[SYCL] Move kernel_compiler related information

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/graph_impl.cpp

@@ -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/helpers.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();