Fix E2E test for kernel compiler

Author: uditagarwal97

sycl/source/detail/persistent_device_code_cache.cpp

@@ -54,8 +54,7 @@ LockCacheItem::~LockCacheItem() {
 }
 
 // Returns true if the specified format is either SPIRV or a native binary.
-static bool
-IsSupportedImageFormat(ur::DeviceBinaryType Format) {
+static bool IsSupportedImageFormat(ur::DeviceBinaryType Format) {
   return Format == SYCL_DEVICE_BINARY_TYPE_SPIRV ||
          Format == SYCL_DEVICE_BINARY_TYPE_NATIVE;
 }
@@ -210,6 +209,16 @@ void PersistentDeviceCodeCache::repopulateCacheSizeFile(
   const std::string CacheSizeFileName = "cache_size.txt";
   const std::string CacheSizeFile = CacheRoot + "/" + CacheSizeFileName;
 
+  // Create cache root, if it does not exist.
+  try {
+    if (!OSUtil::isPathPresent(CacheRoot))
+      OSUtil::makeDir(CacheRoot.c_str());
+  } catch (...) {
+    throw sycl::exception(make_error_code(errc::runtime),
+                          "Failed to create cache root directory: " +
+                              CacheRoot);
+  }
+
   // If the cache size file is not present, calculate the size of the cache size
   // directory and write it to the file.
   if (!OSUtil::isPathPresent(CacheSizeFile)) {

sycl/test-e2e/KernelCompiler/kernel_compiler_cache_eviction.cpp

@@ -6,32 +6,20 @@
 //
 //===----------------------------------------------------------------------===//
 
+// Tests on-disk cache and eviction with kernel_compiler.
+
 // REQUIRES: ocloc && (opencl || level_zero)
 // UNSUPPORTED: accelerator
 
 // -- Test the kernel_compiler with OpenCL source.
 // RUN: %{build} -o %t.out
-// RUN: %{run} %t.out
-// RUN: %{l0_leak_check} %{run} %t.out
 
 // -- Test again, with caching.
-// DEFINE: %{cache_vars} = env SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_TRACE=5 SYCL_CACHE_DIR=%t/cache_dir
-// RUN: rm -rf %t/cache_dir
-// RUN:  %{cache_vars} %t.out 2>&1 |  FileCheck %s --check-prefixes=CHECK-WRITTEN-TO-CACHE
-// RUN:  %{cache_vars} %t.out 2>&1 |  FileCheck %s --check-prefixes=CHECK-READ-FROM-CACHE
-
-// -- Add leak check.
+// DEFINE: %{cache_vars} = env SYCL_CACHE_PERSISTENT=1 SYCL_CACHE_TRACE=7 SYCL_CACHE_DIR=%t/cache_dir SYCL_CACHE_MAX_SIZE=23000
 // RUN: rm -rf %t/cache_dir
-// RUN:  %{l0_leak_check} %{cache_vars} %t.out 2>&1 |  FileCheck %s --check-prefixes=CHECK-WRITTEN-TO-CACHE
-// RUN:  %{l0_leak_check} %{cache_vars} %t.out 2>&1 |  FileCheck %s --check-prefixes=CHECK-READ-FROM-CACHE
+// RUN: %{cache_vars} %t.out 2>&1 | FileCheck %s --check-prefix=CHECK
 
-// CHECK-WRITTEN-TO-CACHE: [Persistent Cache]: enabled
-// CHECK-WRITTEN-TO-CACHE-NOT: [kernel_compiler Persistent Cache]: using cached binary
-// CHECK-WRITTEN-TO-CACHE: [kernel_compiler Persistent Cache]: binary has been cached
-
-// CHECK-READ-FROM-CACHE: [Persistent Cache]: enabled
-// CHECK-READ-FROM-CACHE-NOT: [kernel_compiler Persistent Cache]: binary has been cached
-// CHECK-READ-FROM-CACHE: [kernel_compiler Persistent Cache]: using cached binary
+// CHECK: [Persistent Cache]: enabled
 
 #include <sycl/detail/core.hpp>
 #include <sycl/kernel_bundle.hpp>
@@ -47,54 +35,11 @@ __kernel void her_kernel(__global int *in, __global int *out) {
 }
 )===";
 
-auto constexpr BadCLSource = R"===(
-__kernel void my_kernel(__global int *in, __global int *out) {
-  size_t i = get_global_id(0) +  no semi-colon!!
-  out[i] = in[i]*2 + 100;
-}
-)===";
-/*
-Compile Log:
-1:3:34: error: use of undeclared identifier 'no'
-  size_t i = get_global_id(0) +  no semi-colon!!
-                                 ^
-1:3:36: error: expected ';' at end of declaration
-  size_t i = get_global_id(0) +  no semi-colon!!
-                                   ^
-                                   ;
-
-Build failed with error code: -11
-
-=============
-
-*/
-
 using namespace sycl;
 
-void testSyclKernel(sycl::queue &Q, sycl::kernel Kernel, int multiplier,
-                    int added) {
-  constexpr int N = 4;
-  cl_int InputArray[N] = {0, 1, 2, 3};
-  cl_int OutputArray[N] = {};
-
-  sycl::buffer InputBuf(InputArray, sycl::range<1>(N));
-  sycl::buffer OutputBuf(OutputArray, sycl::range<1>(N));
-
-  Q.submit([&](sycl::handler &CGH) {
-    CGH.set_arg(0, InputBuf.get_access<sycl::access::mode::read>(CGH));
-    CGH.set_arg(1, OutputBuf.get_access<sycl::access::mode::write>(CGH));
-    CGH.parallel_for(sycl::range<1>{N}, Kernel);
-  });
-
-  sycl::host_accessor Out{OutputBuf};
-  for (int I = 0; I < N; I++)
-    assert(Out[I] == ((I * multiplier) + added));
-}
-
 void test_build_and_run() {
   namespace syclex = sycl::ext::oneapi::experimental;
   using source_kb = sycl::kernel_bundle<sycl::bundle_state::ext_oneapi_source>;
-  using exe_kb = sycl::kernel_bundle<sycl::bundle_state::executable>;
 
   // only one device is supported at this time, so we limit the queue and
   // context to that
@@ -112,69 +57,66 @@ void test_build_and_run() {
     return;
   }
 
+  auto CreateAndVerifyKB = [](source_kb &kbSrc,
+                              std::vector<std::string> &&BuildFlags) {
+    std::string log;
+    std::vector<sycl::device> devs = kbSrc.get_devices();
+    sycl::context ctxRes = kbSrc.get_context();
+    sycl::backend beRes = kbSrc.get_backend();
+
+    auto kb =
+        syclex::build(kbSrc, devs,
+                      syclex::properties{syclex::build_options{BuildFlags},
+                                         syclex::save_log{&log}});
+
+    bool hasMyKernel = kb.ext_oneapi_has_kernel("my_kernel");
+    bool hasHerKernel = kb.ext_oneapi_has_kernel("her_kernel");
+    bool notExistKernel = kb.ext_oneapi_has_kernel("not_exist");
+    assert(hasMyKernel && "my_kernel should exist, but doesn't");
+    assert(hasHerKernel && "her_kernel should exist, but doesn't");
+    assert(!notExistKernel && "non-existing kernel should NOT exist.");
+  };
+
   source_kb kbSrc = syclex::create_kernel_bundle_from_source(
       ctx, syclex::source_language::opencl, CLSource);
-  // compilation of empty prop list, no devices
-  exe_kb kbExe1 = syclex::build(kbSrc);
 
   // compilation with props and devices
-  std::string log;
   std::vector<std::string> flags{"-cl-fast-relaxed-math",
                                  "-cl-finite-math-only"};
-  std::vector<sycl::device> devs = kbSrc.get_devices();
-  sycl::context ctxRes = kbSrc.get_context();
-  assert(ctxRes == ctx);
-  sycl::backend beRes = kbSrc.get_backend();
-  assert(beRes == ctx.get_backend());
-
-  exe_kb kbExe2 = syclex::build(
-      kbSrc, devs,
-      syclex::properties{syclex::build_options{flags}, syclex::save_log{&log}});
-
-  bool hasMyKernel = kbExe2.ext_oneapi_has_kernel("my_kernel");
-  bool hasHerKernel = kbExe2.ext_oneapi_has_kernel("her_kernel");
-  bool notExistKernel = kbExe2.ext_oneapi_has_kernel("not_exist");
-  assert(hasMyKernel && "my_kernel should exist, but doesn't");
-  assert(hasHerKernel && "her_kernel should exist, but doesn't");
-  assert(!notExistKernel && "non-existing kernel should NOT exist, but does?");
-
-  sycl::kernel my_kernel = kbExe2.ext_oneapi_get_kernel("my_kernel");
-  sycl::kernel her_kernel = kbExe2.ext_oneapi_get_kernel("her_kernel");
-
-  auto my_num_args = my_kernel.get_info<sycl::info::kernel::num_args>();
-  assert(my_num_args == 2 && "my_kernel should take 2 args");
-
-  testSyclKernel(q, my_kernel, 2, 100);
-  testSyclKernel(q, her_kernel, 5, 1000);
-}
 
-void test_error() {
-  namespace syclex = sycl::ext::oneapi::experimental;
-  using source_kb = sycl::kernel_bundle<sycl::bundle_state::ext_oneapi_source>;
-  using exe_kb = sycl::kernel_bundle<sycl::bundle_state::executable>;
-
-  // only one device is supported at this time, so we limit the queue and
-  // context to that
-  sycl::device d{sycl::default_selector_v};
-  sycl::context ctx{d};
-  sycl::queue q{ctx, d};
-
-  bool ok =
-      q.get_device().ext_oneapi_can_compile(syclex::source_language::opencl);
-  if (!ok) {
-    return;
-  }
-
-  try {
-    source_kb kbSrc = syclex::create_kernel_bundle_from_source(
-        ctx, syclex::source_language::opencl, BadCLSource);
-    exe_kb kbExe1 = syclex::build(kbSrc);
-    assert(false && "we should not be here.");
-  } catch (sycl::exception &e) {
-    // nice!
-    assert(e.code() == sycl::errc::build);
-  }
-  // any other error will escape and cause the test to fail ( as it should ).
+  // Device image #1
+  // CHECK: [Persistent Cache]: Cache size file not present. Creating one.
+  // CHECK-NEXT: [Persistent Cache]: Cache size file created.
+  // CHECK-NEXT: [kernel_compiler Persistent Cache]: binary has been cached: [[DEVIMG1:.*]]
+  // CHECK-NEXT: [Persistent Cache]: Updating the cache size file.
+  CreateAndVerifyKB(kbSrc, {});
+
+  // Device image #2
+  // CHECK-NEXT: [kernel_compiler Persistent Cache]: binary has been cached: [[DEVIMG2:.*]]
+  // CHECK-NEXT: [Persistent Cache]: Updating the cache size file.
+  CreateAndVerifyKB(kbSrc, {flags[0]});
+
+  // Device image #3
+  // CHECK: [kernel_compiler Persistent Cache]: binary has been cached: [[DEVIMG3:.*]]
+  // CHECK: [Persistent Cache]: Updating the cache size file.
+  CreateAndVerifyKB(kbSrc, {flags[1]});
+
+  // Re-insert device image #1
+  // CHECK: [kernel_compiler Persistent Cache]: using cached binary: [[DEVIMG1]]
+  CreateAndVerifyKB(kbSrc, {});
+
+  // Device image #4
+  // CHECK: [kernel_compiler Persistent Cache]: binary has been cached: [[DEVIMG4:.*]]
+  // CHECK: [Persistent Cache]: Updating the cache size file.
+  // CHECK: [Persistent Cache]: Cache eviction triggered.
+  // CHECK: [Persistent Cache]: File removed: [[DEVIMG2]]
+  // CHECK: [Persistent Cache]: File removed: [[DEVIMG3]]
+  // CHECK: [Persistent Cache]: File removed: [[DEVIMG1]]
+  CreateAndVerifyKB(kbSrc, {flags[0], flags[1]});
+
+  // Re-insert device image #4
+  // CHECK: [kernel_compiler Persistent Cache]: using cached binary: [[DEVIMG4]]
+  CreateAndVerifyKB(kbSrc, {flags[0], flags[1]});
 }
 
 int main() {
@@ -184,7 +126,6 @@ int main() {
 
 #ifdef SYCL_EXT_ONEAPI_KERNEL_COMPILER
   test_build_and_run();
-  test_error();
 #else
   static_assert(false, "Kernel Compiler feature test macro undefined");
 #endif