Cache found service kernels

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

Author: steffenlarsen

sycl/source/detail/program_manager/program_manager.cpp

@@ -64,16 +64,6 @@ ProgramManager &ProgramManager::getInstance() {
   return GlobalHandler::instance().getProgramManager();
 }
 
-// This function allows for identifying "service" kernels. A SYCL service kernel
-// is a kernel that has not been defined by the user but is instead generated by
-// the SYCL runtime. Service kernel name types must be declared in the
-// sycl::detail::__sycl_service_kernel__ namespace which is exclusively used for
-// this purpose. As such service kernels can be identified by
-// __sycl_service_kernel__ appearing in the mangled kernel name.
-static bool isServiceKernel(const std::string &KernelName) {
-  return KernelName.find("__sycl_service_kernel__") != std::string::npos;
-}
-
 static RT::PiProgram
 createBinaryProgram(const ContextImplPtr Context, const device &Device,
                     const unsigned char *Data, size_t DataLen,
@@ -1055,9 +1045,14 @@ void ProgramManager::addImages(pi_device_binaries DeviceBinary) {
           (void)Result;
           assert(Result.second && "Kernel sets are not disjoint");
 
-          // Skip creating unique kernel ID if it is a service kernel
-          if (isServiceKernel(EntriesIt->name))
+          // Skip creating unique kernel ID if it is a service kernel.
+          // SYCL service kernels are identified by having
+          // __sycl_service_kernel__ in the mangled name, primarily as part of
+          // the namespace of the name type.
+          if (std::strstr(EntriesIt->name, "__sycl_service_kernel__")) {
+            m_ServiceKernels.insert(EntriesIt->name);
             continue;
+          }
 
           // ... and create a unique kernel ID for the entry
           std::shared_ptr<detail::kernel_id_impl> KernelIDImpl =
@@ -1351,7 +1346,8 @@ ProgramManager::getSYCLDeviceImagesWithCompatibleState(
 
           if (KernelID == m_KernelIDs.end()) {
             // Service kernels do not have kernel IDs
-            assert(isServiceKernel(EntriesIt->name) &&
+            assert(m_ServiceKernels.find(EntriesIt->name) !=
+                       m_ServiceKernels.end() &&
                    "Kernel ID in device binary missing from cache");
             continue;
           }

sycl/source/detail/program_manager/program_manager.hpp

@@ -22,6 +22,7 @@
 #include <map>
 #include <memory>
 #include <unordered_map>
+#include <unordered_set>
 #include <vector>
 
 // +++ Entry points referenced by the offload wrapper object {
@@ -284,6 +285,13 @@ class ProgramManager {
   /// \ref Sync::getGlobalLock() while holding this mutex.
   std::mutex m_KernelIDsMutex;
 
+  /// Caches all found service kernels to expedite future checks. A SYCL service
+  /// kernel is a kernel that has not been defined by the user but is instead
+  /// generated by the SYCL runtime. Service kernel name types must be declared
+  /// in the sycl::detail::__sycl_service_kernel__ namespace which is
+  /// exclusively used for this purpose.
+  std::unordered_set<std::string> m_ServiceKernels;
+
   // Keeps track of pi_program to image correspondence. Needed for:
   // - knowing which specialization constants are used in the program and
   //   injecting their current values before compiling the SPIR-V; the binary