[SYCL] Drop kernel name type backward compatibility

sycl/include/sycl/handler.hpp

@@ -19,7 +19,6 @@
 #include <sycl/detail/impl_utils.hpp>
 #include <sycl/detail/kernel_desc.hpp>
 #include <sycl/detail/kernel_launch_helper.hpp>
-#include <sycl/detail/kernel_name_str_t.hpp>
 #include <sycl/detail/reduction_forward.hpp>
 #include <sycl/detail/string.hpp>
 #include <sycl/detail/string_view.hpp>
@@ -530,7 +529,7 @@ class __SYCL_EXPORT handler {
 
 #ifndef __INTEL_PREVIEW_BREAKING_CHANGES
   /// \return a string containing name of SYCL kernel.
-  detail::ABINeutralKernelNameStrT getKernelName();
+  detail::string_view getKernelName();
 
   template <typename LambdaNameT> bool lambdaAndKernelHaveEqualName() {
     // TODO It is unclear a kernel and a lambda/functor must to be equal or not
@@ -541,7 +540,7 @@ class __SYCL_EXPORT handler {
     assert(MKernel && "MKernel is not initialized");
     constexpr std::string_view LambdaName =
         detail::CompileTimeKernelInfo<LambdaNameT>.Name;
-    detail::ABINeutralKernelNameStrT KernelName = getKernelName();
+    detail::string_view KernelName = getKernelName();
     return KernelName == LambdaName;
   }
 #endif
@@ -3151,7 +3150,7 @@ class __SYCL_EXPORT handler {
 #endif
   std::vector<detail::LocalAccessorImplPtr> MLocalAccStorage;
   std::vector<std::shared_ptr<detail::stream_impl>> MStreamStorage;
-  detail::ABINeutralKernelNameStrT MKernelName;
+  detail::string_view MKernelName;
   /// Storage for a sycl::kernel object.
   std::shared_ptr<detail::kernel_impl> MKernel;
   /// Pointer to the source host memory or accessor(depending on command type).

sycl/source/detail/device_global_map.hpp

@@ -9,13 +9,13 @@
 #pragma once
 
 #include <mutex>
+#include <string_view>
 #include <unordered_map>
 
 #include <detail/compiler.hpp>
 #include <detail/device_binary_image.hpp>
 #include <detail/device_global_map_entry.hpp>
 #include <sycl/detail/defines_elementary.hpp>
-#include <sycl/detail/kernel_name_str_t.hpp>
 #include <sycl/kernel_bundle.hpp>
 
 namespace sycl {
@@ -161,7 +161,7 @@ class DeviceGlobalMap {
 
   size_t size() const { return MDeviceGlobals.size(); }
 
-  size_t count(const KernelNameStrT &UniqueId) const {
+  size_t count(const std::string_view &UniqueId) const {
     return MDeviceGlobals.count(UniqueId);
   }
 
@@ -173,7 +173,7 @@ class DeviceGlobalMap {
   bool MOwnerControlledCleanup = true;
 
   // Maps between device_global identifiers and associated information.
-  std::unordered_map<KernelNameStrT, std::unique_ptr<DeviceGlobalMapEntry>>
+  std::unordered_map<std::string_view, std::unique_ptr<DeviceGlobalMapEntry>>
       MDeviceGlobals;
   std::unordered_map<const void *, DeviceGlobalMapEntry *> MPtr2DeviceGlobal;
 

sycl/source/detail/device_image_impl.cpp

@@ -32,7 +32,7 @@ std::shared_ptr<kernel_impl> device_image_impl::tryGetExtensionKernel(
 
       auto UrProgram = get_ur_program();
       auto [UrKernel, CacheMutex, ArgMask] =
-          PM.getOrCreateKernel(Context, KernelNameStrT(AdjustedName),
+          PM.getOrCreateKernel(Context, AdjustedName,
                                /*PropList=*/{}, UrProgram);
       return std::make_shared<kernel_impl>(
           std::move(UrKernel), *getSyclObjImpl(Context), shared_from_this(),

sycl/source/detail/device_kernel_info.cpp

@@ -13,16 +13,11 @@ inline namespace _V1 {
 namespace detail {
 
 DeviceKernelInfo::DeviceKernelInfo(const CompileTimeKernelInfoTy &Info)
-    : CompileTimeKernelInfoTy(Info)
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-      ,
-      Name(Info.Name.data())
-#endif
-{
+    : CompileTimeKernelInfoTy(Info) {
   init(Name.data());
 }
 
-void DeviceKernelInfo::init(KernelNameStrRefT KernelName) {
+void DeviceKernelInfo::init(std::string_view KernelName) {
   auto &PM = detail::ProgramManager::getInstance();
   MImplicitLocalArgPos = PM.kernelImplicitLocalArgPos(KernelName);
 #ifndef __INTEL_PREVIEW_BREAKING_CHANGES

sycl/source/detail/device_kernel_info.hpp

@@ -11,12 +11,12 @@
 #include <detail/kernel_arg_mask.hpp>
 #include <hash_table8.hpp>
 #include <sycl/detail/compile_time_kernel_info.hpp>
-#include <sycl/detail/kernel_name_str_t.hpp>
 #include <sycl/detail/spinlock.hpp>
 #include <sycl/detail/ur.hpp>
 
 #include <mutex>
 #include <optional>
+#include <string_view>
 
 namespace sycl {
 inline namespace _V1 {
@@ -89,18 +89,12 @@ struct FastKernelSubcacheT {
 // into this structure and get rid of the other KernelName -> * maps.
 class DeviceKernelInfo : public CompileTimeKernelInfoTy {
 public:
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-  // Needs to own the kernel name string in non-preview builds since we pass it
-  // using a temporary string instead of a string view there.
-  std::string Name;
-#endif
-
 #ifndef __INTEL_PREVIEW_BREAKING_CHANGES
   DeviceKernelInfo() = default;
 #endif
   DeviceKernelInfo(const CompileTimeKernelInfoTy &Info);
 
-  void init(KernelNameStrRefT KernelName);
+  void init(std::string_view KernelName);
 #ifndef __INTEL_PREVIEW_BREAKING_CHANGES
   // Initialize default-created entry that has no data recorded:
   void initIfEmpty(const CompileTimeKernelInfoTy &Info);

sycl/source/detail/handler_impl.hpp

@@ -61,9 +61,7 @@ class handler_impl {
            HandlerSubmissionState::EXPLICIT_KERNEL_BUNDLE_STATE;
   }
 
-  KernelNameStrRefT getKernelName() const {
-    return MKernelData.getKernelName();
-  }
+  std::string_view getKernelName() const { return MKernelData.getKernelName(); }
 
   /// Registers mutually exclusive submission states.
   HandlerSubmissionState MSubmissionState = HandlerSubmissionState::NO_STATE;

sycl/source/detail/helpers.cpp

@@ -35,7 +35,7 @@ __SYCL_EXPORT void waitEvents(std::vector<sycl::event> DepEvents) {
 #endif
 
 const RTDeviceBinaryImage *retrieveKernelBinary(queue_impl &Queue,
-                                                KernelNameStrRefT KernelName,
+                                                std::string_view KernelName,
                                                 CGExecKernel *KernelCG) {
   device_impl &Dev = Queue.getDeviceImpl();
   bool isNvidia = Dev.getBackend() == backend::ext_oneapi_cuda;

sycl/source/detail/helpers.hpp

@@ -9,7 +9,6 @@
 #pragma once
 
 #include <sycl/detail/impl_utils.hpp>
-#include <sycl/detail/kernel_name_str_t.hpp>
 #include <sycl/detail/type_traits.hpp>
 
 #include <ur_api.h>
@@ -18,6 +17,7 @@
 #include <iterator>
 #include <memory>
 #include <queue>
+#include <string_view>
 #include <tuple>
 #include <variant>
 #include <vector>
@@ -32,7 +32,7 @@ class queue_impl;
 class RTDeviceBinaryImage;
 
 const RTDeviceBinaryImage *
-retrieveKernelBinary(queue_impl &Queue, KernelNameStrRefT KernelName,
+retrieveKernelBinary(queue_impl &Queue, std::string_view KernelName,
                      CGExecKernel *CGKernel = nullptr);
 
 template <typename SyclTy, typename... Iterators> class variadic_iterator {

sycl/source/detail/jit_compiler.cpp

@@ -144,7 +144,7 @@ static ::jit_compiler::BinaryFormat getTargetFormat(queue_impl &Queue) {
 
 ur_kernel_handle_t jit_compiler::materializeSpecConstants(
     queue_impl &Queue, const RTDeviceBinaryImage *BinImage,
-    KernelNameStrRefT KernelName,
+    std::string_view KernelName,
     const std::vector<unsigned char> &SpecConstBlob) {
 #ifndef _WIN32
   if (!BinImage) {

sycl/source/detail/jit_compiler.hpp

@@ -10,7 +10,6 @@
 
 #include <detail/jit_device_binaries.hpp>
 #include <detail/queue_impl.hpp>
-#include <sycl/detail/kernel_name_str_t.hpp>
 #include <sycl/feature_test.hpp>
 #if SYCL_EXT_JIT_ENABLE
 #include <JITBinaryInfo.h>
@@ -21,6 +20,7 @@
 #include <functional>
 #include <memory>
 #include <mutex>
+#include <string_view>
 #include <unordered_map>
 
 namespace jit_compiler {
@@ -41,7 +41,7 @@ class jit_compiler {
   ur_kernel_handle_t
   materializeSpecConstants(queue_impl &Queue,
                            const RTDeviceBinaryImage *BinImage,
-                           KernelNameStrRefT KernelName,
+                           std::string_view KernelName,
                            const std::vector<unsigned char> &SpecConstBlob);
 
   std::pair<sycl_device_binaries, std::string> compileSYCL(

sycl/source/detail/kernel_bundle_impl.hpp

@@ -18,7 +18,6 @@
 #include <sycl/backend_types.hpp>
 #include <sycl/context.hpp>
 #include <sycl/detail/common.hpp>
-#include <sycl/detail/kernel_name_str_t.hpp>
 #include <sycl/device.hpp>
 #include <sycl/kernel_bundle.hpp>
 
@@ -27,6 +26,7 @@
 #include <cstdint>
 #include <cstring>
 #include <memory>
+#include <string_view>
 #include <unordered_set>
 #include <vector>
 
@@ -987,8 +987,7 @@ class kernel_bundle_impl
         std::move(SelectedImage), *this, ArgMask, UrProgram, CacheMutex);
   }
 
-  std::shared_ptr<kernel_impl>
-  tryGetKernel(detail::KernelNameStrRefT Name) const {
+  std::shared_ptr<kernel_impl> tryGetKernel(std::string_view Name) const {
     // TODO: For source-based kernels, it may be faster to keep a map between
     //       {kernel_name, device} and their corresponding image.
     // First look through the kernels registered in source-based images.

sycl/source/detail/kernel_data.hpp

@@ -306,9 +306,9 @@ class KernelData {
         Kprop.get<cuda::cluster_size_key<3>>()->MProperty);
   }
 
-  KernelNameStrRefT getKernelName() const {
+  std::string_view getKernelName() const {
     assert(MDeviceKernelInfoPtr);
-    return static_cast<KernelNameStrRefT>(MDeviceKernelInfoPtr->Name);
+    return MDeviceKernelInfoPtr->Name;
   }
 
   void processArg(void *Ptr, const detail::kernel_param_kind_t &Kind,

sycl/source/detail/kernel_id_impl.hpp

@@ -8,7 +8,11 @@
 
 #pragma once
 
-#include <sycl/detail/kernel_name_str_t.hpp>
+#include <sycl/kernel_bundle.hpp>
+
+#include <cstring>
+#include <string>
+#include <string_view>
 
 namespace sycl {
 inline namespace _V1 {
@@ -33,7 +37,7 @@ struct EqualByNameComp {
 // identificator
 class kernel_id_impl {
 public:
-  kernel_id_impl(KernelNameStrT Name) : MName(std::move(Name)) {}
+  kernel_id_impl(std::string_view Name) : MName(std::move(Name)) {}
   kernel_id_impl(){};
   const char *get_name() { return MName.data(); }
 

sycl/source/detail/kernel_impl.hpp

@@ -243,7 +243,7 @@ class kernel_impl {
     return MIsInterop
                ? MInteropDeviceKernelInfo
                : ProgramManager::getInstance().getOrCreateDeviceKernelInfo(
-                     KernelNameStrT(getName()));
+                     std::string_view(getName()));
   }
 
 private:

sycl/source/detail/kernel_program_cache.hpp

@@ -15,7 +15,6 @@
 #include <detail/platform_impl.hpp>
 #include <detail/unordered_multimap.hpp>
 #include <sycl/detail/common.hpp>
-#include <sycl/detail/kernel_name_str_t.hpp>
 #include <sycl/detail/locked.hpp>
 #include <sycl/detail/os_util.hpp>
 #include <sycl/detail/spinlock.hpp>
@@ -215,7 +214,7 @@ class KernelProgramCache {
 
   using KernelCacheT = emhash8::HashMap<
       ur_program_handle_t,
-      emhash8::HashMap<KernelNameStrT, std::shared_ptr<KernelBuildResult>>>;
+      emhash8::HashMap<std::string_view, std::shared_ptr<KernelBuildResult>>>;
 
   class FastKernelSubcacheWrapper {
   public:
@@ -260,7 +259,7 @@ class KernelProgramCache {
   };
 
   using FastKernelCacheT =
-      emhash8::HashMap<KernelNameStrT, FastKernelSubcacheWrapper>;
+      emhash8::HashMap<std::string_view, FastKernelSubcacheWrapper>;
 
   // DS to hold data and functions related to Program cache eviction.
   struct EvictionList {
@@ -365,7 +364,7 @@ class KernelProgramCache {
   // set.
   template <typename MsgType>
   static inline void traceKernel(const MsgType &Msg,
-                                 KernelNameStrRefT KernelName,
+                                 std::string_view KernelName,
                                  bool IsFastKernelCache = false) {
     if (!SYCLConfig<SYCL_CACHE_TRACE>::isTraceInMemCache())
       return;
@@ -430,7 +429,7 @@ class KernelProgramCache {
   }
 
   std::pair<std::shared_ptr<KernelBuildResult>, bool>
-  getOrInsertKernel(ur_program_handle_t Program, KernelNameStrRefT KernelName) {
+  getOrInsertKernel(ur_program_handle_t Program, std::string_view KernelName) {
     auto LockedCache = acquireKernelsPerProgramCache();
     auto &Cache = LockedCache.get()[Program];
     auto [It, DidInsert] = Cache.try_emplace(KernelName, nullptr);
@@ -443,7 +442,7 @@ class KernelProgramCache {
   }
 
   FastKernelCacheValPtr
-  tryToGetKernelFast(KernelNameStrRefT KernelName, ur_device_handle_t Device,
+  tryToGetKernelFast(std::string_view KernelName, ur_device_handle_t Device,
                      FastKernelSubcacheT &KernelSubcache) {
     const FastKernelSubcacheEntriesT &SubcacheEntries = KernelSubcache.Entries;
     FastKernelSubcacheReadLockT SubcacheLock{KernelSubcache.Mutex};
@@ -462,7 +461,7 @@ class KernelProgramCache {
     return FastKernelCacheValPtr();
   }
 
-  void saveKernel(KernelNameStrRefT KernelName, ur_device_handle_t Device,
+  void saveKernel(std::string_view KernelName, ur_device_handle_t Device,
                   const FastKernelCacheValPtr &CacheVal,
                   FastKernelSubcacheT &KernelSubcache) {
     if (SYCLConfig<SYCL_IN_MEM_CACHE_EVICTION_THRESHOLD>::
@@ -842,8 +841,9 @@ class KernelProgramCache {
 
   // Map between fast kernel cache keys and program handle.
   // MFastKernelCacheMutex will be used for synchronization.
-  std::unordered_map<ur_program_handle_t,
-                     std::vector<std::pair<KernelNameStrT, ur_device_handle_t>>>
+  std::unordered_map<
+      ur_program_handle_t,
+      std::vector<std::pair<std::string_view, ur_device_handle_t>>>
       MProgramToFastKernelCacheKeyMap;
 
   EvictionList MEvictionList;