[SYCL] Move kernel_compiler related information (#17380)

This commit moves the kernel compiler related information from the
kernel bundles to the device images. This separation allows the
implementation to properly join and (in the near future) link kernel
bundles created from different paths.

---------

Signed-off-by: Larsen, Steffen <[email protected]>

Author: steffenlarsen

sycl/source/backend.cpp

@@ -299,7 +299,8 @@ make_kernel_bundle(ur_native_handle_t NativeHandle,
   // symbols (e.g. when kernel_bundle is supposed to be joined with another).
   auto KernelIDs = std::make_shared<std::vector<kernel_id>>();
   auto DevImgImpl = std::make_shared<device_image_impl>(
-      nullptr, TargetContext, Devices, State, KernelIDs, UrProgram);
+      nullptr, TargetContext, Devices, State, KernelIDs, UrProgram,
+      ImageOriginInterop);
   device_image_plain DevImg{DevImgImpl};
 
   return std::make_shared<kernel_bundle_impl>(TargetContext, Devices, DevImg);

sycl/source/detail/device_image_impl.hpp

@@ -1534,18 +1534,10 @@ void exec_graph_impl::populateURKernelUpdateStructs(
   std::shared_ptr<sycl::detail::kernel_impl> SyclKernelImpl = nullptr;
   const sycl::detail::KernelArgMask *EliminatedArgMask = nullptr;
 
-  // Use kernel_bundle if available unless it is interop.
-  // Interop bundles can't be used in the first branch, because the kernels
-  // in interop kernel bundles (if any) do not have kernel_id
-  // and can therefore not be looked up, but since they are self-contained
-  // they can simply be launched directly.
-  if (KernelBundleImplPtr && !KernelBundleImplPtr->isInterop()) {
-    const auto &KernelName = ExecCG.MKernelName;
-    kernel_id KernelID =
-        sycl::detail::ProgramManager::getInstance().getSYCLKernelID(KernelName);
-    kernel SyclKernel =
-        KernelBundleImplPtr->get_kernel(KernelID, KernelBundleImplPtr);
-    SyclKernelImpl = sycl::detail::getSyclObjImpl(SyclKernel);
+  if (auto SyclKernelImpl = KernelBundleImplPtr
+                                ? KernelBundleImplPtr->tryGetKernel(
+                                      ExecCG.MKernelName, KernelBundleImplPtr)
+                                : std::shared_ptr<kernel_impl>{nullptr}) {
     UrKernel = SyclKernelImpl->getHandleRef();
     EliminatedArgMask = SyclKernelImpl->getKernelArgMask();
   } else if (Kernel != nullptr) {

sycl/source/detail/graph_impl.cpp

@@ -72,17 +72,14 @@ retrieveKernelBinary(const QueueImplPtr &Queue, const char *KernelName,
 
   const RTDeviceBinaryImage *DeviceImage = nullptr;
   ur_program_handle_t Program = nullptr;
-  if (KernelCG->getKernelBundle() != nullptr) {
+  auto KernelBundleImpl = KernelCG->getKernelBundle();
+  if (auto SyclKernelImpl =
+          KernelBundleImpl
+              ? KernelBundleImpl->tryGetKernel(KernelName, KernelBundleImpl)
+              : std::shared_ptr<kernel_impl>{nullptr}) {
     // Retrieve the device image from the kernel bundle.
-    auto KernelBundle = KernelCG->getKernelBundle();
-    kernel_id KernelID =
-        detail::ProgramManager::getInstance().getSYCLKernelID(KernelName);
-
-    auto SyclKernel = detail::getSyclObjImpl(
-        KernelBundle->get_kernel(KernelID, KernelBundle));
-
-    DeviceImage = SyclKernel->getDeviceImage()->get_bin_image_ref();
-    Program = SyclKernel->getDeviceImage()->get_ur_program_ref();
+    DeviceImage = SyclKernelImpl->getDeviceImage()->get_bin_image_ref();
+    Program = SyclKernelImpl->getDeviceImage()->get_ur_program_ref();
   } else if (KernelCG->MSyclKernel != nullptr) {
     DeviceImage = KernelCG->MSyclKernel->getDeviceImage()->get_bin_image_ref();
     Program = KernelCG->MSyclKernel->getDeviceImage()->get_ur_program_ref();