return back setDeviceKernelInfoPtr

Author: vinser52

sycl/include/sycl/handler.hpp

@@ -851,8 +851,8 @@ class __SYCL_EXPORT handler {
     if constexpr (KernelHasName) {
       // TODO support ESIMD in no-integration-header case too.
 
-      setKernelInfo((void *)MHostKernel->getPtr(),
-                    &detail::getDeviceKernelInfo<KernelName>());
+      setDeviceKernelInfoPtr(&detail::getDeviceKernelInfo<KernelName>());
+      setKernelInfo((void *)MHostKernel->getPtr());
 
       constexpr std::string_view KernelNameStr =
           detail::getKernelName<KernelName>();
@@ -3659,8 +3659,7 @@ class __SYCL_EXPORT handler {
                      detail::kernel_param_desc_t (*KernelParamDescGetter)(int),
                      bool KernelIsESIMD, bool KernelHasSpecialCaptures);
 #endif
-  void setKernelInfo(void *KernelFuncPtr,
-                     detail::DeviceKernelInfo *DeviceKernelInfoPtr);
+  void setKernelInfo(void *KernelFuncPtr);
 
   void instantiateKernelOnHost(void *InstantiateKernelOnHostPtr);
 
@@ -3688,6 +3687,7 @@ class __SYCL_EXPORT handler {
   void setKernelNameBasedCachePtr(
       detail::KernelNameBasedCacheT *KernelNameBasedCachePtr);
 #endif
+  void setDeviceKernelInfoPtr(detail::DeviceKernelInfo *DeviceKernelInfoPtr);
 
   queue getQueue();
 

sycl/source/detail/kernel_data.hpp

@@ -107,11 +107,7 @@ class KernelData {
   }
 #endif
 
-  void setKernelInfo(void *KernelFuncPtr,
-                     detail::DeviceKernelInfo *DeviceKernelInfoPtr) {
-    MKernelFuncPtr = KernelFuncPtr;
-    MDeviceKernelInfoPtr = DeviceKernelInfoPtr;
-  }
+  void setKernelInfo(void *KernelFuncPtr) { MKernelFuncPtr = KernelFuncPtr; }
 
   bool usesAssert() const { return MDeviceKernelInfoPtr->usesAssert(); }
 

sycl/source/handler.cpp

@@ -2329,9 +2329,14 @@ void handler::setKernelInfo(
 }
 #endif
 
-void handler::setKernelInfo(void *KernelFuncPtr,
-                            detail::DeviceKernelInfo *DeviceKernelInfoPtr) {
-  impl->MKernelData.setKernelInfo(KernelFuncPtr, DeviceKernelInfoPtr);
+void handler::setDeviceKernelInfoPtr(
+    sycl::detail::DeviceKernelInfo *DeviceKernelInfoPtr) {
+  assert(!impl->MKernelData.getDeviceKernelInfoPtr() && "Already set!");
+  impl->MKernelData.setDeviceKernelInfoPtr(DeviceKernelInfoPtr);
+}
+
+void handler::setKernelInfo(void *KernelFuncPtr) {
+  impl->MKernelData.setKernelInfo(KernelFuncPtr);
 }
 
 void handler::instantiateKernelOnHost(void *InstantiateKernelOnHostPtr) {

sycl/test/abi/sycl_symbols_linux.dump

@@ -3545,8 +3545,8 @@ _ZN4sycl3_V17handler12addReductionERKSt10shared_ptrIKvE
 _ZN4sycl3_V17handler12setArgHelperEiONS0_6streamE
 _ZN4sycl3_V17handler12setArgHelperEiRNS0_6detail22work_group_memory_implE
 _ZN4sycl3_V17handler13getKernelNameEv
+_ZN4sycl3_V17handler13setKernelInfoEPv
 _ZN4sycl3_V17handler13setKernelInfoEPviPFNS0_6detail19kernel_param_desc_tEiEbb
-_ZN4sycl3_V17handler13setKernelInfoEPvPNS0_6detail16DeviceKernelInfoE
 _ZN4sycl3_V17handler14addAccessorReqESt10shared_ptrINS0_6detail16AccessorImplHostEE
 _ZN4sycl3_V17handler14setNDRangeUsedEb
 _ZN4sycl3_V17handler15ext_oneapi_copyENS0_3ext6oneapi12experimental16image_mem_handleENS0_5rangeILi3EEERKNS4_16image_descriptorEPvS7_S7_S7_
@@ -3593,6 +3593,7 @@ _ZN4sycl3_V17handler21setKernelWorkGroupMemEm
 _ZN4sycl3_V17handler21setUserFacingNodeTypeENS0_3ext6oneapi12experimental9node_typeE
 _ZN4sycl3_V17handler22ext_oneapi_fill2d_implEPvmPKvmmm
 _ZN4sycl3_V17handler22memcpyFromDeviceGlobalEPvPKvbmm
+_ZN4sycl3_V17handler22setDeviceKernelInfoPtrEPNS0_6detail16DeviceKernelInfoE
 _ZN4sycl3_V17handler22setHandlerKernelBundleENS0_6kernelE
 _ZN4sycl3_V17handler22setHandlerKernelBundleERKSt10shared_ptrINS0_6detail18kernel_bundle_implEE
 _ZN4sycl3_V17handler22setKernelClusterLaunchENS0_5rangeILi1EEE

sycl/test/abi/sycl_symbols_windows.dump

@@ -4409,6 +4409,7 @@
 ?setArgsHelper@handler@_V1@sycl@@AEAAXH@Z
 ?setArgsToAssociatedAccessors@handler@_V1@sycl@@AEAAXXZ
 ?setDevice@HostProfilingInfo@detail@_V1@sycl@@QEAAXPEAVdevice_impl@234@@Z
+?setDeviceKernelInfoPtr@handler@_V1@sycl@@AEAAXPEAVDeviceKernelInfo@detail@23@@Z
 ?setHandlerKernelBundle@handler@_V1@sycl@@AEAAXAEBV?$shared_ptr@Vkernel_bundle_impl@detail@_V1@sycl@@@std@@@Z
 ?setHandlerKernelBundle@handler@_V1@sycl@@AEAAXVkernel@23@@Z
 ?setKernelCacheConfig@handler@_V1@sycl@@AEAAXW4StableKernelCacheConfig@123@@Z
@@ -4417,6 +4418,7 @@
 ?setKernelClusterLaunch@handler@_V1@sycl@@AEAAXV?$range@$02@23@@Z
 ?setKernelClusterLaunch@handler@_V1@sycl@@AEAAXV?$range@$02@23@H@Z
 ?setKernelInfo@handler@_V1@sycl@@AEAAXPEAXHP6A?AUkernel_param_desc_t@detail@23@H@Z_N2@Z
+?setKernelInfo@handler@_V1@sycl@@AEAAXPEAX@Z
 ?setKernelIsCooperative@handler@_V1@sycl@@AEAAX_N@Z
 ?setKernelNameBasedCachePtr@handler@_V1@sycl@@AEAAXPEAUKernelNameBasedCacheT@detail@23@@Z
 ?setKernelWorkGroupMem@handler@_V1@sycl@@AEAAX_K@Z