[SYCL][E2E] fixed UR FileCheck conditions to match new traces

makes llvm work with this change: oneapi-src/unified-runtime#2101 fixes this bug: oneapi-src/unified-runtime#2002
lslusarczyk · Sep 20, 2024 · b72f1a1 · b72f1a1
1 parent fb2fcc2
commit b72f1a1
Show file tree

Hide file tree

Showing 33 changed files with 183 additions and 183 deletions.
diff --git a/sycl/cmake/modules/FetchUnifiedRuntime.cmake b/sycl/cmake/modules/FetchUnifiedRuntime.cmake
@@ -116,14 +116,14 @@ if(SYCL_UR_USE_FETCH_CONTENT)
  CACHE PATH "Path to external '${name}' adapter source dir" FORCE)
  endfunction()
 
- set(UNIFIED_RUNTIME_REPO "https://github.com/oneapi-src/unified-runtime.git")
- # commit 897bcfb10a7e13b8e413278550e845322f6bd560 (HEAD, origin/main, origin/HEAD)
- # Merge: 18514924 71cdf253
- # Author: Omar Ahmed <omar.ahmed@codeplay.com>
- # Date: Thu Sep 19 14:35:55 2024 +0100
- # Merge pull request #2022 from nrspruit/add_3_channel_image_support
- # [L0] Add 3-channel image support
- set(UNIFIED_RUNTIME_TAG 897bcfb10a7e13b8e413278550e845322f6bd560)
+ set(UNIFIED_RUNTIME_REPO "https://github.com/lslusarczyk/unified-runtime.git")
+ # commit 185149248dd257bd37482aac43307a136204c051
+ # Merge: 2af159d4 d619bcd1
+ # Author: Piotr Balcer <piotr.balcer@intel.com>
+ # Date: Thu Sep 19 11:02:27 2024 +0200
+ #  Merge pull request #1934 from yingcong-wu/yc/0806-exclude-shadow-from-coredump
+ #  [DeviceSanitizer] Exclude shadow memory from coredump file for CPU device.
+ set(UNIFIED_RUNTIME_TAG 2b5d716653915ebab91e1f92e3ff02cca5f1a572)
 
  set(UMF_BUILD_EXAMPLES OFF CACHE INTERNAL "EXAMPLES")
  # Due to the use of dependentloadflag and no installer for UMF and hwloc we need

diff --git a/sycl/test-e2e/Basic/buffer/native_buffer_creation_flags.cpp b/sycl/test-e2e/Basic/buffer/native_buffer_creation_flags.cpp
@@ -20,7 +20,7 @@ int main() {
  Q.submit([&](handler &Cgh) {
  // Now that we have a read-write host allocation, check that the native
  // buffer is created with the UR_MEM_FLAG_USE_HOST_POINTER flag.
- // CHECK: urMemBufferCreate
+ // CHECK: ---> urMemBufferCreate
  // CHECK-SAME: UR_MEM_FLAG_USE_HOST_POINTER
  auto BufAcc = Buf.get_access<access::mode::read>(Cgh);
  Cgh.single_task<Foo>([=]() { int A = BufAcc[0]; });

diff --git a/sycl/test-e2e/Basic/buffer/subbuffer_overlap.cpp b/sycl/test-e2e/Basic/buffer/subbuffer_overlap.cpp
@@ -16,13 +16,13 @@ int main() {
  for (auto &e : sycl::host_accessor{b})
  e = idx++ % size;
 
- // CHECK: urMemBufferPartition
+ // CHECK: ---> urMemBufferPartition
  // CHECK: .origin = 256, .size = 64
  q.submit([&](sycl::handler &cgh) {
  sycl::accessor acc{sub1, cgh};
  cgh.parallel_for(size, [=](auto id) { acc[id] += 1; });
  });
- // CHECK: urMemBufferPartition
+ // CHECK: ---> urMemBufferPartition
  // CHECK: .origin = 256, .size = 128
  q.submit([&](sycl::handler &cgh) {
  sycl::accessor acc{sub2, cgh};

diff --git a/sycl/test-e2e/Basic/event_release.cpp b/sycl/test-e2e/Basic/event_release.cpp
@@ -29,8 +29,8 @@ int main() {
  // Buffer destruction triggers execution graph cleanup, check that both
  // events (one for launching the kernel and one for memory transfer to host)
  // are released.
- // CHECK: urEventRelease
- // CHECK: urEventRelease
+ // CHECK: ---> urEventRelease
+ // CHECK: ---> urEventRelease
  assert(Val == Gold);
  // CHECK: End of main scope
  std::cout << "End of main scope" << std::endl;

diff --git a/sycl/test-e2e/Basic/fill_accessor_ur.cpp b/sycl/test-e2e/Basic/fill_accessor_ur.cpp
@@ -129,19 +129,19 @@ int main() {
 }
 
 // CHECK: start testFill_Buffer1D
-// CHECK: urEnqueueMemBufferFill
+// CHECK: ---> urEnqueueMemBufferFill
 // CHECK: start testFill_Buffer1D -- OFFSET
-// CHECK: urEnqueueMemBufferFill
+// CHECK: ---> urEnqueueMemBufferFill
 
 // CHECK: start testFill_Buffer2D
-// CHECK: urEnqueueMemBufferFill
+// CHECK: ---> urEnqueueMemBufferFill
 // CHECK: start testFill_Buffer2D -- OFFSET
-// CHECK: urEnqueueKernelLaunch
+// CHECK: ---> urEnqueueKernelLaunch
 
 // CHECK: start testFill_Buffer3D
-// CHECK: urEnqueueMemBufferFill
+// CHECK: ---> urEnqueueMemBufferFill
 // CHECK: start testFill_Buffer3D -- OFFSET
-// CHECK: urEnqueueKernelLaunch
+// CHECK: ---> urEnqueueKernelLaunch
 
 // CHECK: start testFill_ZeroDim
-// CHECK: urEnqueueMemBufferFill
+// CHECK: ---> urEnqueueMemBufferFill
diff --git a/sycl/test-e2e/Basic/kernel_bundle/kernel_bundle_api.cpp b/sycl/test-e2e/Basic/kernel_bundle/kernel_bundle_api.cpp
@@ -150,26 +150,26 @@ int main() {
 
  sycl::kernel_bundle<sycl::bundle_state::object> KernelBundleObject1 =
  sycl::compile(KernelBundleInput1, KernelBundleInput1.get_devices());
- // CHECK:---> urProgramCreate
+ // CHECK:<--- urProgramCreate
  // CHECK-SAME:, .phProgram = {{.*}} ([[PROGRAM_HANDLE1:[0-9a-fA-Fx]+]])
  // CHECK-SAME: -> UR_RESULT_SUCCESS;
  //
- // CHECK:---> urProgramCompile
+ // CHECK:<--- urProgramCompile
  // CHECK-SAME: .hProgram = [[PROGRAM_HANDLE1]]
 
  sycl::kernel_bundle<sycl::bundle_state::object> KernelBundleObject2 =
  sycl::compile(KernelBundleInput2, KernelBundleInput2.get_devices());
- // CHECK:---> urProgramCreate
+ // CHECK:<--- urProgramCreate
  // CHECK-SAME:, .phProgram = {{.*}} ([[PROGRAM_HANDLE2:[0-9a-fA-Fx]+]])
  // CHECK-SAME: -> UR_RESULT_SUCCESS;
  //
- // CHECK:---> urProgramCompile(
+ // CHECK:<--- urProgramCompile(
  // CHECK-SAME: .hProgram = [[PROGRAM_HANDLE2]]
 
  sycl::kernel_bundle<sycl::bundle_state::executable> KernelBundleExecutable =
  sycl::link({KernelBundleObject1, KernelBundleObject2},
  KernelBundleObject1.get_devices());
- // CHECK:---> urProgramLink{{.*}} -> UR_RESULT_SUCCESS;
+ // CHECK:<--- urProgramLink{{.*}} -> UR_RESULT_SUCCESS;
  // UR tracing doesn't allow checking for all input programs so far.
 
  assert(KernelBundleExecutable.has_kernel(Kernel1ID));
@@ -179,14 +179,14 @@ int main() {
  KernelBundleExecutable2 =
  sycl::build(KernelBundleInput1, KernelBundleInput1.get_devices());
 
- // CHECK:---> urProgramCreate
+ // CHECK:<--- urProgramCreate
  // CHECK-SAME:, .phProgram = {{.*}} ([[PROGRAM_HANDLE3:[0-9a-fA-Fx]+]])
  // CHECK-SAME: -> UR_RESULT_SUCCESS;
  //
- // CHECK:---> urProgramBuild(
+ // CHECK:<--- urProgramBuild(
  // CHECK-SAME: .hProgram = [[PROGRAM_HANDLE3]]
  //
- // CHECK:---> urProgramRetain(
+ // CHECK:<--- urProgramRetain(
  // CHECK-SAME: .hProgram = [[PROGRAM_HANDLE3]]
  // CHECK-SAME:-> UR_RESULT_SUCCESS;
 
@@ -204,31 +204,31 @@ int main() {
  sycl::kernel_bundle KernelBundleExecutable =
  sycl::get_kernel_bundle<sycl::bundle_state::executable>(Ctx, {Dev},
  {Kernel3ID});
- // CHECK:---> urProgramCreate
+ // CHECK:<--- urProgramCreate
  // CHECK-SAME:, .phProgram = {{.*}} ([[PROGRAM_HANDLE4:[0-9a-fA-Fx]+]])
  // CHECK-SAME: -> UR_RESULT_SUCCESS;
  //
- // CHECK:---> urProgramBuild(
+ // CHECK:<--- urProgramBuild(
  // CHECK-SAME: .hProgram = [[PROGRAM_HANDLE4]]
  //
- // CHECK:---> urProgramRetain(
+ // CHECK:<--- urProgramRetain(
  // CHECK-SAME: .hProgram = [[PROGRAM_HANDLE4]]
  // CHECK-SAME:-> UR_RESULT_SUCCESS;
  //
- // CHECK:---> urKernelCreate(
+ // CHECK:<--- urKernelCreate(
  // CHECK-SAME: .hProgram = [[PROGRAM_HANDLE4]]
  // CHECK-SAME: .pKernelName = {{[0-9a-fA-Fx]+}} (_ZTS11Kernel3Name)
  // CHECK-SAME: .phKernel = {{[0-9a-fA-Fx]+}} ([[KERNEL_HANDLE:[0-9a-fA-Fx]+]])
  // CHECK-SAME: -> UR_RESULT_SUCCESS;
  //
- // CHECK:---> urKernelRetain(
+ // CHECK:<--- urKernelRetain(
  // CHECK-SAME: .hKernel = [[KERNEL_HANDLE]]
  // CHECK-SAME:-> UR_RESULT_SUCCESS;
  //
- // CHECK:---> urEnqueueKernelLaunch(
+ // CHECK:<--- urEnqueueKernelLaunch(
  // CHECK-SAME: .hKernel = [[KERNEL_HANDLE]]
  //
- // CHECK:---> urKernelRelease(
+ // CHECK:<--- urKernelRelease(
  // CHECK-SAME: .hKernel = [[KERNEL_HANDLE]]
  // CHECK-SAME:-> UR_RESULT_SUCCESS;
 

diff --git a/sycl/test-e2e/DeviceCodeSplit/grf.cpp b/sycl/test-e2e/DeviceCodeSplit/grf.cpp
@@ -145,10 +145,10 @@ int main(void) {
 
 // CHECK: ---> urKernelCreate({{.*}}SingleGRF{{.*}}-> UR_RESULT_SUCCESS
 
-// CHECK-NO-VAR: urProgramBuild{{.*}}-ze-opt-large-register-file
-// CHECK-WITH-VAR: urProgramBuild{{.*}}-g -ze-opt-large-register-file
-// CHECK-AUTO-NO-VAR: urProgramBuild{{.*}}-ze-intel-enable-auto-large-GRF-mode
-// CHECK-AUTO-WITH-VAR: urProgramBuild{{.*}}-g -ze-intel-enable-auto-large-GRF-mode
+// CHECK-NO-VAR: ---> urProgramBuild{{.*}}-ze-opt-large-register-file
+// CHECK-WITH-VAR: ---> urProgramBuild{{.*}}-g -ze-opt-large-register-file
+// CHECK-AUTO-NO-VAR: ---> urProgramBuild{{.*}}-ze-intel-enable-auto-large-GRF-mode
+// CHECK-AUTO-WITH-VAR: ---> urProgramBuild{{.*}}-g -ze-intel-enable-auto-large-GRF-mode
 // CHECK-SAME: -> UR_RESULT_SUCCESS
 
 // CHECK: ---> urKernelCreate({{.*}}SpecifiedGRF{{.*}}-> UR_RESULT_SUCCESS
diff --git a/sycl/test-e2e/DiscardEvents/discard_events_usm.cpp b/sycl/test-e2e/DiscardEvents/discard_events_usm.cpp
@@ -71,7 +71,7 @@
 // CHECK: ---> urEnqueueUSMPrefetch
 // CHECK: .phEvent = nullptr
 //
-// CHECK: ---> urEnqueueUSMAdvise
+// CHECK: <--- urEnqueueUSMAdvise
 // CHECK: .phEvent = nullptr
 // CHECK-SAME: ) -> {{UR_RESULT_SUCCESS|UR_RESULT_ERROR_ADAPTER_SPECIFIC}}
 //

diff --git a/sycl/test-e2e/ESIMD/esimd_check_vc_codegen.cpp b/sycl/test-e2e/ESIMD/esimd_check_vc_codegen.cpp
@@ -94,4 +94,4 @@ int main(void) {
 // Don't use -NEXT here to split the line because we need to allow for the
 // possbility of a BuildExp( that fails with UNSUPPORTED followed by a Build(
 // that results in SUCCESS
-// CHECK: ---> urProgramBuild{{(Exp)?}}({{.*}}-vc-codegen{{.*}} -> UR_RESULT_SUCCESS
+// CHECK: <--- urProgramBuild{{(Exp)?}}({{.*}}-vc-codegen{{.*}} -> UR_RESULT_SUCCESS
diff --git a/sycl/test-e2e/Graph/Explicit/kernel_bundle.cpp b/sycl/test-e2e/Graph/Explicit/kernel_bundle.cpp
@@ -5,27 +5,27 @@
 // Checks the UR call trace to ensure that the bundle kernel of the single task
 // is used.
 
-// CHECK:---> urProgramCreateWithIL(
+// CHECK:<--- urProgramCreateWithIL(
 // CHECK-SAME: .phProgram = {{.*}} ([[PROGRAM_HANDLE1:[0-9a-fA-Fx]+]])
 
 //
-// CHECK:---> urProgramBuildExp(
+// CHECK:<--- urProgramBuildExp(
 // CHECK-SAME: .hProgram = [[PROGRAM_HANDLE1]]
 //
-// CHECK:---> urProgramRetain(.hProgram = [[PROGRAM_HANDLE1]]) -> UR_RESULT_SUCCESS
+// CHECK:<--- urProgramRetain(.hProgram = [[PROGRAM_HANDLE1]]) -> UR_RESULT_SUCCESS
 
-// CHECK:---> urKernelCreate(
+// CHECK:<--- urKernelCreate(
 // CHECK-SAME: .hProgram = [[PROGRAM_HANDLE1]]
 // CHECK-SAME: .pKernelName = {{.*}} (_ZTS11Kernel1Name)
 // CHECK-SAME: .phKernel = {{.*}} ([[KERNEL_HANDLE:[0-9a-fA-Fx]+]])
 // CHECK-SAME: -> UR_RESULT_SUCCESS
 //
-// CHECK:---> urKernelRetain(.hKernel = [[KERNEL_HANDLE]]) -> UR_RESULT_SUCCESS
+// CHECK:<--- urKernelRetain(.hKernel = [[KERNEL_HANDLE]]) -> UR_RESULT_SUCCESS
 //
-// CHECK:---> urCommandBufferAppendKernelLaunchExp(
+// CHECK:<--- urCommandBufferAppendKernelLaunchExp(
 // CHECK-SAME: .hKernel = [[KERNEL_HANDLE]]
 //
-// CHECK:---> urKernelRelease(.hKernel = [[KERNEL_HANDLE]]) -> UR_RESULT_SUCCESS
+// CHECK:<--- urKernelRelease(.hKernel = [[KERNEL_HANDLE]]) -> UR_RESULT_SUCCESS
 
 #define GRAPH_E2E_EXPLICIT
 

diff --git a/sycl/test-e2e/Graph/Explicit/memadvise.cpp b/sycl/test-e2e/Graph/Explicit/memadvise.cpp
@@ -10,14 +10,14 @@
 // impact results but only performances, we verify
 // that a node is correctly added by checking UR function calls.
 
-// CHECK: urCommandBufferAppendUSMAdviseExp
+// CHECK: ---> urCommandBufferAppendUSMAdviseExp
 // CHECK-SAME: .hCommandBuffer = 0x[[#%x,COMMAND_BUFFER:]]
 // CHECK-SAME: .pMemory = 0x[[#%x,PTR:]]
 // CHECK-SAME: .size = 400
 // CHECK-SAME: .pSyncPoint = {{.*}} (0x[[#%x,ADVISE_SYNC_POINT:]])
 // CHECK-SAME: -> UR_RESULT_SUCCESS
 
-// CHECK: urCommandBufferAppendKernelLaunchExp(
+// CHECK: ---> urCommandBufferAppendKernelLaunchExp(
 // CHECK-SAME: .hCommandBuffer = 0x[[#%x,COMMAND_BUFFER]]
 // CHECK-SAME: .hKernel = 0x[[#%x,KERNEL:]]
 // CHECK-SAME: .workDim = 1

diff --git a/sycl/test-e2e/Graph/Explicit/prefetch.cpp b/sycl/test-e2e/Graph/Explicit/prefetch.cpp
@@ -10,13 +10,13 @@
 // impact results but only performances, we verify
 // that a node is correctly added by checking UR function calls
 
-// CHECK: urCommandBufferAppendUSMPrefetchExp(
+// CHECK: ---> urCommandBufferAppendUSMPrefetchExp(
 // CHECK-SAME: .hCommandBuffer = 0x[[#%x,COMMAND_BUFFER:]]
 // CHECK-SAME: .pMemory = 0x[[#%x,PTR:]]
 // CHECK-SAME: .size = 400
 // CHECK-SAME: .pSyncPoint = {{.*}} (0x[[#%x,PREFETCH_SYNC_POINT:]])
 
-// CHECK: urCommandBufferAppendKernelLaunchExp(
+// CHECK: ---> urCommandBufferAppendKernelLaunchExp(
 // CHECK-SAME: .hCommandBuffer = 0x[[#%x,COMMAND_BUFFER]]
 // CHECK-SAME: .hKernel = 0x[[#%x,KERNEL:]]
 // CHECK-SAME: .workDim = 1

diff --git a/sycl/test-e2e/Graph/RecordReplay/memadvise.cpp b/sycl/test-e2e/Graph/RecordReplay/memadvise.cpp
@@ -10,13 +10,13 @@
 // impact results but only performances, we verify
 // that a node is correctly added by checking UR function calls.
 
-// CHECK: urCommandBufferAppendUSMAdviseExp
+// CHECK: ---> urCommandBufferAppendUSMAdviseExp
 // CHECK-SAME: .hCommandBuffer = 0x[[#%x,COMMAND_BUFFER:]]
 // CHECK-SAME: .pMemory = 0x[[#%x,PTR:]]
 // CHECK-SAME: .size = 400
 // CHECK-SAME: .pSyncPoint = {{.*}} (0x[[#%x,ADVISE_SYNC_POINT:]])
 
-// CHECK: urCommandBufferAppendKernelLaunchExp(
+// CHECK: ---> urCommandBufferAppendKernelLaunchExp(
 // CHECK-SAME: .hCommandBuffer = 0x[[#%x,COMMAND_BUFFER]]
 // CHECK-SAME: .hKernel = 0x[[#%x,KERNEL:]]
 // CHECK-SAME: .workDim = 1

diff --git a/sycl/test-e2e/Graph/RecordReplay/prefetch.cpp b/sycl/test-e2e/Graph/RecordReplay/prefetch.cpp
@@ -10,13 +10,13 @@
 // impact results but only performances, we verify
 // that a node is correctly added by checking UR function calls
 
-// CHECK: urCommandBufferAppendUSMPrefetchExp(
+// CHECK: ---> urCommandBufferAppendUSMPrefetchExp(
 // CHECK-SAME: .hCommandBuffer = 0x[[#%x,COMMAND_BUFFER:]]
 // CHECK-SAME: .pMemory = 0x[[#%x,PTR:]]
 // CHECK-SAME: .size = 400
 // CHECK-SAME: .pSyncPoint = {{.*}} (0x[[#%x,PREFETCH_SYNC_POINT:]])
 
-// CHECK: urCommandBufferAppendKernelLaunchExp(
+// CHECK: ---> urCommandBufferAppendKernelLaunchExp(
 // CHECK-SAME: .hCommandBuffer = 0x[[#%x,COMMAND_BUFFER]]
 // CHECK-SAME: .hKernel = 0x[[#%x,KERNEL:]]
 // CHECK-SAME: .workDim = 1

diff --git a/sycl/test-e2e/KernelAndProgram/cache_env_vars.cpp b/sycl/test-e2e/KernelAndProgram/cache_env_vars.cpp
@@ -25,12 +25,12 @@
 // Some backends will call urProgramBuild and some will call
 // urProgramBuildExp depending on urProgramBuildExp support.
 
-// CHECK-BUILD-NOT: urProgramCreateWithBinary(
-// CHECK-BUILD: urProgramCreateWithIL(
-// CHECK-BUILD: urProgramBuild{{(Exp)?}}(
+// CHECK-BUILD-NOT: ---> urProgramCreateWithBinary(
+// CHECK-BUILD: ---> urProgramCreateWithIL(
+// CHECK-BUILD: ---> urProgramBuild{{(Exp)?}}(
 
-// CHECK-CACHE-NOT: urProgramCreateWithIL(
-// CHECK-CACHE: urProgramCreateWithBinary(
-// CHECK-CACHE: urProgramBuild{{(Exp)?}}(
+// CHECK-CACHE-NOT: ---> urProgramCreateWithIL(
+// CHECK-CACHE: ---> urProgramCreateWithBinary(
+// CHECK-CACHE: ---> urProgramBuild{{(Exp)?}}(
 
 #include "cache_env_vars.hpp"
diff --git a/sycl/test-e2e/KernelAndProgram/cache_env_vars_lin.cpp b/sycl/test-e2e/KernelAndProgram/cache_env_vars_lin.cpp
@@ -24,12 +24,12 @@
 
 // Some backends will call urProgramBuild and some will call urProgramBuildExp depending on urProgramBuildExp support.
 
-// CHECK-BUILD-NOT: urProgramCreateWithBinary(
-// CHECK-BUILD: urProgramCreateWithIL(
-// CHECK-BUILD: urProgramBuild{{(Exp)?}}(
+// CHECK-BUILD-NOT: ---> urProgramCreateWithBinary(
+// CHECK-BUILD: ---> urProgramCreateWithIL(
+// CHECK-BUILD: ---> urProgramBuild{{(Exp)?}}(
 
-// CHECK-CACHE-NOT: urProgramCreateWithIL(
-// CHECK-CACHE: urProgramCreateWithBinary(
-// CHECK-CACHE: urProgramBuild{{(Exp)?}}(
+// CHECK-CACHE-NOT: ---> urProgramCreateWithIL(
+// CHECK-CACHE: ---> urProgramCreateWithBinary(
+// CHECK-CACHE: ---> urProgramBuild{{(Exp)?}}(
 
 #include "cache_env_vars.hpp"
diff --git a/sycl/test-e2e/KernelAndProgram/cache_env_vars_win.cpp b/sycl/test-e2e/KernelAndProgram/cache_env_vars_win.cpp
@@ -19,12 +19,12 @@
 // RUN: env SYCL_CACHE_PERSISTENT=1 AppData=%t/cache_dir SYCL_UR_TRACE=2 env -u SYCL_CACHE_DIR %t.out | FileCheck %s --check-prefixes=CHECK-BUILD
 // RUN: env SYCL_CACHE_PERSISTENT=1 AppData=%t/cache_dir SYCL_UR_TRACE=2 env -u SYCL_CACHE_DIR %t.out | FileCheck %s --check-prefixes=CHECK-CACHE
 
-// CHECK-BUILD-NOT: urProgramCreateWithBinary(
-// CHECK-BUILD: urProgramCreateWithIL(
-// CHECK-BUILD: urProgramBuild{{(Exp)?}}(
+// CHECK-BUILD-NOT: ---> urProgramCreateWithBinary(
+// CHECK-BUILD: ---> urProgramCreateWithIL(
+// CHECK-BUILD: ---> urProgramBuild{{(Exp)?}}(
 
-// CHECK-CACHE-NOT: urProgramCreateWithIL(
-// CHECK-CACHE: urProgramCreateWithBinary(
-// CHECK-CACHE: urProgramBuild{{(Exp)?}}(
+// CHECK-CACHE-NOT: ---> urProgramCreateWithIL(
+// CHECK-CACHE: ---> urProgramCreateWithBinary(
+// CHECK-CACHE: ---> urProgramBuild{{(Exp)?}}(
 
 #include "cache_env_vars.hpp"