[SYCL] Fix DeviceCodeSplit checks to eliminate KernelInfo usage (#9316)

SYCL implementation does not properly support KernelInfo usage in
multiple translation units. In this test integration header with
File1Kern1 is included to main source file where we execute that kernel.
In the second file get_kernel_id is used for checks but integration
header is not attached to that file since we do not execute kernel there
and we got UB because trying to use get_name() function of KI. This
problem is a known one. This test doesn't target to verify this behavior
- it is targeted to verify that kernels from different files are added to
different device images if we built it with related option. A different
check is implemented that does not involve UB case and enables initially
disabled checks because of the problem above.

Signed-off-by: Tikhomirova, Kseniya <[email protected]>

Author: KseniyaTikhomirova

sycl/test-e2e/DeviceCodeSplit/Inputs/split-per-source-second-file.cpp

@@ -6,14 +6,13 @@ void runKernelsFromFile2() {
   {
     sycl::buffer<int, 1> Buf(&Data, sycl::range<1>(1));
     auto KernelID1 = sycl::get_kernel_id<File2Kern1>();
-    auto KernelID2 = sycl::get_kernel_id<File1Kern1>();
-    auto KernelID3 = sycl::get_kernel_id<File1Kern2>();
     auto KB = sycl::get_kernel_bundle<sycl::bundle_state::executable>(
         Q.get_context(), {KernelID1});
     auto Krn = KB.get_kernel(KernelID1);
 
-    assert(!KB.has_kernel(KernelID2));
-    assert(!KB.has_kernel(KernelID3));
+    std::vector<sycl::kernel_id> KernelIDStorage = KB.get_kernel_ids();
+    assert(KernelIDStorage.size() == 1);
+    assert(KernelIDStorage[0] == KernelID1);
 
     Q.submit([&](sycl::handler &Cgh) {
       auto Acc = Buf.get_access<sycl::access::mode::read_write>(Cgh);

sycl/test-e2e/DeviceCodeSplit/split-per-source-main.cpp

@@ -11,41 +11,40 @@
 int main() {
   sycl::queue Q;
   int Data = 0;
-  {
-    sycl::buffer<int, 1> Buf(&Data, sycl::range<1>(1));
-    auto KernelID = sycl::get_kernel_id<File1Kern1>();
-    auto KB = sycl::get_kernel_bundle<sycl::bundle_state::executable>(
-        Q.get_context(), {KernelID});
-    auto Krn = KB.get_kernel(KernelID);
 
-    assert(KB.has_kernel(KernelID));
-    // TODO uncomment once the KernelInfo in multiple translation units
-    // bug is fixed.
-    // assert(!Prg.has_kernel<File2Kern1>());
+  auto KernelID = sycl::get_kernel_id<File1Kern1>();
+  auto KB = sycl::get_kernel_bundle<sycl::bundle_state::executable>(
+      Q.get_context(), {KernelID});
+  assert(KB.has_kernel(KernelID));
+  auto Krn1 = KB.get_kernel(KernelID);
+
+  auto KernelID2 = sycl::get_kernel_id<File1Kern2>();
+  assert(KB.has_kernel(KernelID2));
+  auto Krn2 = KB.get_kernel(KernelID2);
+
+  std::vector<sycl::kernel_id> KernelIDStorage = KB.get_kernel_ids();
+  assert(KernelIDStorage.size() == 2);
+  assert(std::any_of(
+      KernelIDStorage.begin(), KernelIDStorage.end(),
+      [&KernelID](const sycl::kernel_id &id) { return id == KernelID; }));
+  assert(std::any_of(
+      KernelIDStorage.begin(), KernelIDStorage.end(),
+      [&KernelID2](const sycl::kernel_id &id) { return id == KernelID2; }));
 
+  {
+    sycl::buffer<int, 1> Buf(&Data, sycl::range<1>(1));
     Q.submit([&](sycl::handler &Cgh) {
       auto Acc = Buf.get_access<sycl::access::mode::read_write>(Cgh);
-      Cgh.single_task<File1Kern1>(/*Krn,*/ [=]() { Acc[0] = 1; });
+      Cgh.single_task<File1Kern1>(Krn1, [=]() { Acc[0] = 1; });
     });
   }
   assert(Data == 1);
 
   {
     sycl::buffer<int, 1> Buf(&Data, sycl::range<1>(1));
-    auto KernelID1 = sycl::get_kernel_id<File1Kern1>();
-    auto KernelID2 = sycl::get_kernel_id<File1Kern2>();
-    auto KB = sycl::get_kernel_bundle<sycl::bundle_state::executable>(
-        Q.get_context(), {KernelID1});
-    auto Krn = KB.get_kernel(KernelID2);
-
-    assert(KB.has_kernel(KernelID1));
-    // TODO uncomment once the KernelInfo in multiple translation units
-    // bug is fixed.
-    // assert(!Prg.has_kernel<File2Kern1>());
-
     Q.submit([&](sycl::handler &Cgh) {
       auto Acc = Buf.get_access<sycl::access::mode::read_write>(Cgh);
-      Cgh.single_task<File1Kern2>(/*Krn,*/ [=]() { Acc[0] = 2; });
+      Cgh.single_task<File1Kern2>(Krn2, [=]() { Acc[0] = 2; });
     });
   }
   assert(Data == 2);