[SYCL][RTC] Implement specified error behavior (intel#17331)

Detect error scenarios in `sycl::ext::oneapi::build` and throw the
errors specified in the extension.

This handles two scenarios:
- Calling build where one of the devices given in the list does not
support the source language of the bundle.
- Calling build where one of the devices given in the list does not
belong to the context of the source bundle.

---------

Signed-off-by: Lukas Sommer <[email protected]>

Author: sommerlukas

sycl/source/detail/kernel_bundle_impl.hpp

@@ -495,7 +495,21 @@ class kernel_bundle_impl {
     std::vector<ur_device_handle_t> DeviceVec;
     DeviceVec.reserve(Devices.size());
     for (const auto &SyclDev : Devices) {
-      ur_device_handle_t Dev = getSyclObjImpl(SyclDev)->getHandleRef();
+      DeviceImplPtr DevImpl = getSyclObjImpl(SyclDev);
+      if (!ContextImpl->hasDevice(DevImpl)) {
+        throw sycl::exception(make_error_code(errc::invalid),
+                              "device not part of kernel_bundle context");
+      }
+      if (!DevImpl->extOneapiCanCompile(MLanguage)) {
+        // This error cannot not be exercised in the current implementation, as
+        // compatibility with a source language depends on the backend's
+        // capabilities and all devices in one context share the same backend in
+        // the current implementation, so this would lead to an error already
+        // during construction of the source bundle.
+        throw sycl::exception(make_error_code(errc::invalid),
+                              "device does not support source language");
+      }
+      ur_device_handle_t Dev = DevImpl->getHandleRef();
       DeviceVec.push_back(Dev);
     }
 

sycl/test-e2e/KernelCompiler/kernel_compiler_basic.cpp

@@ -1,4 +1,4 @@
-//==- kernel_compiler_sycl_jit.cpp --- kernel_compiler extension tests -----==//
+//==---- kernel_compiler_basic.cpp --- kernel_compiler extension tests -----==//
 //
 // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
 // See https://llvm.org/LICENSE.txt for license information.

sycl/test-e2e/KernelCompiler/kernel_compiler_context_error.cpp

@@ -0,0 +1,85 @@
+//==- kernel_compiler_context_error.cpp -- kernel_compiler extension tests -==//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+// REQUIRES: (opencl || level_zero)
+
+// UNSUPPORTED: accelerator
+// UNSUPPORTED-INTENDED: while accelerator is AoT only, this cannot run there.
+
+// RUN: %{build} -o %t.out
+// RUN: %{run-unfiltered-devices} %t.out
+
+#include <sycl/detail/core.hpp>
+#include <sycl/kernel_bundle.hpp>
+#include <sycl/usm.hpp>
+
+namespace syclexp = sycl::ext::oneapi::experimental;
+
+const std::string source = R"""(
+    #include <sycl/sycl.hpp>
+    namespace syclext = sycl::ext::oneapi;
+    namespace syclexp = sycl::ext::oneapi::experimental;
+
+    extern "C"
+    SYCL_EXT_ONEAPI_FUNCTION_PROPERTY((syclexp::nd_range_kernel<1>))
+    void iota(float start, float *ptr) {
+      size_t id = syclext::this_work_item::get_nd_item<1>().get_global_linear_id();
+      ptr[id] = start + static_cast<float>(id);
+    }
+)""";
+
+int main() {
+
+  auto has_multiple_compatible_devices = [](sycl::platform platform) -> bool {
+    auto devices = platform.get_devices();
+    if (devices.size() < 2) {
+      return false;
+    }
+    for (auto dev : devices) {
+      if (!dev.ext_oneapi_can_compile(syclexp::source_language::sycl)) {
+        return false;
+      }
+    }
+    return true;
+  };
+
+  std::vector<sycl::device> all_devices = [&]() -> std::vector<sycl::device> {
+    for (auto platform : sycl::platform::get_platforms()) {
+      if (has_multiple_compatible_devices(platform)) {
+        return platform.get_devices();
+      }
+    }
+    return {};
+  }();
+
+  if (all_devices.size() < 2) {
+    std::cerr << "Cannot find platform with more than 1 device, skipping"
+              << std::endl;
+    return 0;
+  }
+
+  sycl::context single_device_context{all_devices.front()};
+
+  // Create a source kernel bundle with a context that contains only one device.
+  sycl::kernel_bundle<sycl::bundle_state::ext_oneapi_source> kb_src =
+      syclexp::create_kernel_bundle_from_source(
+          single_device_context, syclexp::source_language::sycl, source);
+
+  // Compile the kernel.  There is no need to use the "registered_names"
+  // property because the kernel is declared extern "C".
+  try {
+    syclexp::build(kb_src, all_devices);
+    assert(false && "out-of-context device not detected");
+  } catch (sycl::exception &e) {
+    assert(e.code() == sycl::errc::invalid);
+    assert(std::string(e.what()).find(
+               "device not part of kernel_bundle context") !=
+           std::string::npos);
+  }
+  return 0;
+}

sycl/test/e2e_test_requirements/no_sycl_hpp_in_e2e_tests.cpp

@@ -6,7 +6,7 @@
 // CHECK-DAG: README.md
 // CHECK-DAG: lit.cfg.py
 //
-// CHECK-NUM-MATCHES: 10
+// CHECK-NUM-MATCHES: 11
 //
 // This test verifies that `<sycl/sycl.hpp>` isn't used in E2E tests. Instead,
 // fine-grained includes should used, see