diff --git a/sycl/include/sycl/kernel_bundle.hpp b/sycl/include/sycl/kernel_bundle.hpp index e3d893f25742f..fab864e412480 100644 --- a/sycl/include/sycl/kernel_bundle.hpp +++ b/sycl/include/sycl/kernel_bundle.hpp @@ -186,9 +186,13 @@ class __SYCL_EXPORT kernel_bundle_plain { bool native_specialization_constant() const noexcept; - bool ext_oneapi_has_kernel(const std::string &name); + bool ext_oneapi_has_kernel(const std::string &name) { + return ext_oneapi_has_kernel(detail::string_view{name}); + } - kernel ext_oneapi_get_kernel(const std::string &name); + kernel ext_oneapi_get_kernel(const std::string &name) { + return ext_oneapi_get_kernel(detail::string_view{name}); + } protected: // \returns a kernel object which represents the kernel identified by @@ -214,6 +218,10 @@ class __SYCL_EXPORT kernel_bundle_plain { bool is_specialization_constant_set(const char *SpecName) const noexcept; detail::KernelBundleImplPtr impl; + +private: + bool ext_oneapi_has_kernel(detail::string_view name); + kernel ext_oneapi_get_kernel(detail::string_view name); }; } // namespace detail @@ -897,26 +905,84 @@ __SYCL_EXPORT bool is_source_kernel_bundle_supported(backend BE, source_language Language); __SYCL_EXPORT kernel_bundle +make_kernel_bundle_from_source( + const context &SyclContext, source_language Language, + sycl::detail::string_view Source, + std::vector> + IncludePairsVec); + +inline kernel_bundle make_kernel_bundle_from_source( const context &SyclContext, source_language Language, const std::string &Source, - std::vector> IncludePairsVec); + std::vector> IncludePairsVec) { + size_t n = IncludePairsVec.size(); + std::vector> + PairVec; + PairVec.reserve(n); + for (auto &Pair : IncludePairsVec) + PairVec.push_back({sycl::detail::string_view{Pair.first}, + sycl::detail::string_view{Pair.second}}); + + return make_kernel_bundle_from_source( + SyclContext, Language, sycl::detail::string_view{Source}, PairVec); +} #if (!defined(_HAS_STD_BYTE) || _HAS_STD_BYTE != 0) __SYCL_EXPORT kernel_bundle make_kernel_bundle_from_source( const context &SyclContext, source_language Language, const std::vector &Bytes, - std::vector> IncludePairsVec); + std::vector> + IncludePairsVec); + +inline kernel_bundle +make_kernel_bundle_from_source( + const context &SyclContext, source_language Language, + const std::vector &Bytes, + std::vector> IncludePairsVec) { + size_t n = IncludePairsVec.size(); + std::vector> + PairVec; + PairVec.reserve(n); + for (auto &Pair : IncludePairsVec) + PairVec.push_back({sycl::detail::string_view{Pair.first}, + sycl::detail::string_view{Pair.second}}); + + return make_kernel_bundle_from_source(SyclContext, Language, Bytes, PairVec); +} #endif -__SYCL_EXPORT kernel_bundle +__SYCL_EXPORT kernel_bundle build_from_source( + kernel_bundle &SourceKB, + const std::vector &Devices, + const std::vector &BuildOptions, + sycl::detail::string *LogPtr, + const std::vector &RegisteredKernelNames); + +inline kernel_bundle build_from_source(kernel_bundle &SourceKB, const std::vector &Devices, const std::vector &BuildOptions, std::string *LogPtr, - const std::vector &RegisteredKernelNames); - + const std::vector &RegisteredKernelNames) { + std::vector Options; + for (const std::string &opt : BuildOptions) + Options.push_back(sycl::detail::string_view{opt}); + + std::vector KernelNames; + for (const std::string &name : RegisteredKernelNames) + KernelNames.push_back(sycl::detail::string_view{name}); + + if (LogPtr) { + sycl::detail::string Log; + auto result = + build_from_source(SourceKB, Devices, Options, &Log, KernelNames); + *LogPtr = Log.c_str(); + return result; + } + return build_from_source(SourceKB, Devices, Options, nullptr, KernelNames); +} } // namespace detail ///////////////////////// diff --git a/sycl/source/kernel_bundle.cpp b/sycl/source/kernel_bundle.cpp index 85a6d0a561bd2..098ff8901cc52 100644 --- a/sycl/source/kernel_bundle.cpp +++ b/sycl/source/kernel_bundle.cpp @@ -115,12 +115,12 @@ bool kernel_bundle_plain::is_specialization_constant_set( return impl->is_specialization_constant_set(SpecName); } -bool kernel_bundle_plain::ext_oneapi_has_kernel(const std::string &name) { - return impl->ext_oneapi_has_kernel(name); +bool kernel_bundle_plain::ext_oneapi_has_kernel(detail::string_view name) { + return impl->ext_oneapi_has_kernel(name.data()); } -kernel kernel_bundle_plain::ext_oneapi_get_kernel(const std::string &name) { - return impl->ext_oneapi_get_kernel(name, impl); +kernel kernel_bundle_plain::ext_oneapi_get_kernel(detail::string_view name) { + return impl->ext_oneapi_get_kernel(name.data(), impl); } ////////////////////////////////// @@ -391,14 +391,24 @@ bool is_source_kernel_bundle_supported(backend BE, source_language Language) { ///////////////////////// using include_pairs_t = std::vector>; - -source_kb make_kernel_bundle_from_source(const context &SyclContext, - source_language Language, - const std::string &Source, - include_pairs_t IncludePairs) { +using include_pairs_view_t = std::vector< + std::pair>; + +source_kb +make_kernel_bundle_from_source(const context &SyclContext, + source_language Language, + sycl::detail::string_view SourceView, + include_pairs_view_t IncludePairViews) { // TODO: if we later support a "reason" why support isn't present // (like a missing shared library etc.) it'd be nice to include it in // the exception message here. + std::string Source{SourceView.data()}; + include_pairs_t IncludePairs; + size_t n = IncludePairViews.size(); + IncludePairs.reserve(n); + for (auto &p : IncludePairViews) + IncludePairs.push_back({p.first.data(), p.second.data()}); + backend BE = SyclContext.get_backend(); if (!is_source_kernel_bundle_supported(BE, Language)) throw sycl::exception(make_error_code(errc::invalid), @@ -417,7 +427,7 @@ source_kb make_kernel_bundle_from_source(const context &SyclContext, source_kb make_kernel_bundle_from_source(const context &SyclContext, source_language Language, const std::vector &Bytes, - include_pairs_t IncludePairs) { + include_pairs_view_t IncludePairs) { (void)IncludePairs; backend BE = SyclContext.get_backend(); if (!is_source_kernel_bundle_supported(BE, Language)) @@ -433,17 +443,32 @@ source_kb make_kernel_bundle_from_source(const context &SyclContext, // syclex::detail::build_from_source(source_kb) => exe_kb ///////////////////////// -exe_kb -build_from_source(source_kb &SourceKB, const std::vector &Devices, - const std::vector &BuildOptions, - std::string *LogPtr, - const std::vector &RegisteredKernelNames) { +exe_kb build_from_source( + source_kb &SourceKB, const std::vector &Devices, + const std::vector &BuildOptions, + sycl::detail::string *LogView, + const std::vector &RegisteredKernelNames) { + std::vector Options; + for (const sycl::detail::string_view option : BuildOptions) + Options.push_back(option.data()); + + std::vector KernelNames; + for (const sycl::detail::string_view name : RegisteredKernelNames) + KernelNames.push_back(name.data()); + + std::string Log; + std::string *LogPtr = nullptr; + if (LogView) + LogPtr = &Log; std::vector UniqueDevices = sycl::detail::removeDuplicateDevices(Devices); std::shared_ptr sourceImpl = getSyclObjImpl(SourceKB); std::shared_ptr KBImpl = sourceImpl->build_from_source( - UniqueDevices, BuildOptions, LogPtr, RegisteredKernelNames); - return sycl::detail::createSyclObjFromImpl(KBImpl); + UniqueDevices, Options, LogPtr, KernelNames); + auto result = sycl::detail::createSyclObjFromImpl(KBImpl); + if (LogView) + *LogView = Log; + return result; } } // namespace detail diff --git a/sycl/test/abi/sycl_abi_neutrality_test.cpp b/sycl/test/abi/sycl_abi_neutrality_test.cpp index 38f7c131bfe52..3b3cb9d48177d 100644 --- a/sycl/test/abi/sycl_abi_neutrality_test.cpp +++ b/sycl/test/abi/sycl_abi_neutrality_test.cpp @@ -18,11 +18,6 @@ // CHECK:_ZN4sycl3_V13ext5intel12experimental15online_compilerILNS3_15source_languageE0EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_ // CHECK:_ZN4sycl3_V13ext5intel12experimental15online_compilerILNS3_15source_languageE1EE7compileIJSt6vectorINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISE_EEEEES8_IhSaIhEERKSE_DpRKT_ // CHECK:_ZN4sycl3_V13ext5intel12experimental9pipe_base13get_pipe_nameB5cxx11EPKv -// CHECK:_ZN4sycl3_V13ext6oneapi12experimental6detail17build_from_sourceERNS0_13kernel_bundleILNS0_12bundle_stateE3EEERKSt6vectorINS0_6deviceESaISA_EERKS9_INSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISK_EEPSK_SO_ -// CHECK:_ZN4sycl3_V13ext6oneapi12experimental6detail30make_kernel_bundle_from_sourceERKNS0_7contextENS3_15source_languageERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESt6vectorISt4pairISE_SE_ESaISJ_EE -// CHECK:_ZN4sycl3_V13ext6oneapi12experimental6detail30make_kernel_bundle_from_sourceERKNS0_7contextENS3_15source_languageERKSt6vectorISt4byteSaISA_EES9_ISt4pairINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESL_ESaISM_EE -// CHECK:_ZN4sycl3_V16detail19kernel_bundle_plain21ext_oneapi_get_kernelERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEE -// CHECK:_ZN4sycl3_V16detail19kernel_bundle_plain21ext_oneapi_has_kernelERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEE // CHECK:_ZN4sycl3_V16detail6OSUtil10getDirNameB5cxx11EPKc // CHECK:_ZN4sycl3_V16detail6OSUtil16getCurrentDSODirB5cxx11Ev // CHECK:_ZN4sycl3_V16opencl13has_extensionERKNS0_6deviceERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEE diff --git a/sycl/test/abi/sycl_symbols_linux.dump b/sycl/test/abi/sycl_symbols_linux.dump index ed9fd1ca2870c..f672d773abbe7 100644 --- a/sycl/test/abi/sycl_symbols_linux.dump +++ b/sycl/test/abi/sycl_symbols_linux.dump @@ -3059,7 +3059,7 @@ _ZN4sycl3_V13ext6oneapi12experimental6detail14image_mem_implC1ERKNS3_16image_des _ZN4sycl3_V13ext6oneapi12experimental6detail14image_mem_implC2ERKNS3_16image_descriptorERKNS0_6deviceERKNS0_7contextE _ZN4sycl3_V13ext6oneapi12experimental6detail14image_mem_implD1Ev _ZN4sycl3_V13ext6oneapi12experimental6detail14image_mem_implD2Ev -_ZN4sycl3_V13ext6oneapi12experimental6detail17build_from_sourceERNS0_13kernel_bundleILNS0_12bundle_stateE3EEERKSt6vectorINS0_6deviceESaISA_EERKS9_INSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESaISK_EEPSK_SO_ +_ZN4sycl3_V13ext6oneapi12experimental6detail17build_from_sourceERNS0_13kernel_bundleILNS0_12bundle_stateE3EEERKSt6vectorINS0_6deviceESaISA_EERKS9_INS0_6detail11string_viewESaISG_EEPNSF_6stringESK_ _ZN4sycl3_V13ext6oneapi12experimental6detail22dynamic_parameter_base11updateValueEPKvm _ZN4sycl3_V13ext6oneapi12experimental6detail22dynamic_parameter_base14updateAccessorEPKNS0_6detail16AccessorBaseHostE _ZN4sycl3_V13ext6oneapi12experimental6detail22dynamic_parameter_baseC1ENS3_13command_graphILNS3_11graph_stateE0EEEmPKv @@ -3083,8 +3083,8 @@ _ZN4sycl3_V13ext6oneapi12experimental6detail24modifiable_command_graphC1ERKNS0_5 _ZN4sycl3_V13ext6oneapi12experimental6detail24modifiable_command_graphC1ERKNS0_7contextERKNS0_6deviceERKNS0_13property_listE _ZN4sycl3_V13ext6oneapi12experimental6detail24modifiable_command_graphC2ERKNS0_5queueERKNS0_13property_listE _ZN4sycl3_V13ext6oneapi12experimental6detail24modifiable_command_graphC2ERKNS0_7contextERKNS0_6deviceERKNS0_13property_listE -_ZN4sycl3_V13ext6oneapi12experimental6detail30make_kernel_bundle_from_sourceERKNS0_7contextENS3_15source_languageERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESt6vectorISt4pairISE_SE_ESaISJ_EE -_ZN4sycl3_V13ext6oneapi12experimental6detail30make_kernel_bundle_from_sourceERKNS0_7contextENS3_15source_languageERKSt6vectorISt4byteSaISA_EES9_ISt4pairINSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEESL_ESaISM_EE +_ZN4sycl3_V13ext6oneapi12experimental6detail30make_kernel_bundle_from_sourceERKNS0_7contextENS3_15source_languageENS0_6detail11string_viewESt6vectorISt4pairISA_SA_ESaISD_EE +_ZN4sycl3_V13ext6oneapi12experimental6detail30make_kernel_bundle_from_sourceERKNS0_7contextENS3_15source_languageERKSt6vectorISt4byteSaISA_EES9_ISt4pairINS0_6detail11string_viewESH_ESaISI_EE _ZN4sycl3_V13ext6oneapi12experimental6detail33is_source_kernel_bundle_supportedENS0_7backendENS3_15source_languageE _ZN4sycl3_V13ext6oneapi12experimental6memcpyENS0_5queueEPvPKvmRKNS0_6detail13code_locationE _ZN4sycl3_V13ext6oneapi12experimental6memsetENS0_5queueEPvimRKNS0_6detail13code_locationE @@ -3256,8 +3256,8 @@ _ZN4sycl3_V16detail18get_kernel_id_implENS1_11string_viewE _ZN4sycl3_V16detail18make_kernel_bundleEmRKNS0_7contextENS0_12bundle_stateENS0_7backendE _ZN4sycl3_V16detail18make_kernel_bundleEmRKNS0_7contextEbNS0_12bundle_stateENS0_7backendE _ZN4sycl3_V16detail18stringifyErrorCodeEi -_ZN4sycl3_V16detail19kernel_bundle_plain21ext_oneapi_get_kernelERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEE -_ZN4sycl3_V16detail19kernel_bundle_plain21ext_oneapi_has_kernelERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEE +_ZN4sycl3_V16detail19kernel_bundle_plain21ext_oneapi_get_kernelENS1_11string_viewE +_ZN4sycl3_V16detail19kernel_bundle_plain21ext_oneapi_has_kernelENS1_11string_viewE _ZN4sycl3_V16detail19kernel_bundle_plain32set_specialization_constant_implEPKcPvm _ZN4sycl3_V16detail20associateWithHandlerERNS0_7handlerEPNS1_16AccessorBaseHostENS0_6access6targetE _ZN4sycl3_V16detail20associateWithHandlerERNS0_7handlerEPNS1_28SampledImageAccessorBaseHostENS0_12image_targetE @@ -3501,6 +3501,7 @@ _ZN4sycl3_V17handler22ext_oneapi_fill2d_implEPvmPKvmmm _ZN4sycl3_V17handler22memcpyFromDeviceGlobalEPvPKvbmm _ZN4sycl3_V17handler22setHandlerKernelBundleENS0_6kernelE _ZN4sycl3_V17handler22setHandlerKernelBundleERKSt10shared_ptrINS0_6detail18kernel_bundle_implEE +_ZN4sycl3_V17handler22setKernelClusterLaunchENS0_5rangeILi3EEEi _ZN4sycl3_V17handler22setKernelIsCooperativeEb _ZN4sycl3_V17handler24GetRangeRoundingSettingsERmS2_S2_ _ZN4sycl3_V17handler24ext_intel_read_host_pipeENS0_6detail11string_viewEPvmb @@ -3508,17 +3509,16 @@ _ZN4sycl3_V17handler24ext_oneapi_memcpy2d_implEPvmPKvmmm _ZN4sycl3_V17handler24ext_oneapi_memset2d_implEPvmimm _ZN4sycl3_V17handler24registerDynamicParameterERNS0_3ext6oneapi12experimental6detail22dynamic_parameter_baseEi _ZN4sycl3_V17handler25ext_intel_write_host_pipeENS0_6detail11string_viewEPvmb -_ZN4sycl3_V17handler26setNDRangeDescriptorPaddedENS0_5rangeILi3EEEbi +_ZN4sycl3_V17handler26associateWithHandlerCommonESt10shared_ptrINS0_6detail16AccessorImplHostEEi _ZN4sycl3_V17handler26setNDRangeDescriptorPaddedENS0_5rangeILi3EEENS0_2idILi3EEEi _ZN4sycl3_V17handler26setNDRangeDescriptorPaddedENS0_5rangeILi3EEES3_NS0_2idILi3EEEi -_ZN4sycl3_V17handler26associateWithHandlerCommonESt10shared_ptrINS0_6detail16AccessorImplHostEEi +_ZN4sycl3_V17handler26setNDRangeDescriptorPaddedENS0_5rangeILi3EEEbi _ZN4sycl3_V17handler27addLifetimeSharedPtrStorageESt10shared_ptrIKvE -_ZN4sycl3_V17handler22setKernelClusterLaunchENS0_5rangeILi3EEEi _ZN4sycl3_V17handler27computeFallbackKernelBoundsEmm _ZN4sycl3_V17handler28extractArgsAndReqsFromLambdaEPcmPKNS0_6detail19kernel_param_desc_tEb _ZN4sycl3_V17handler28memcpyToHostOnlyDeviceGlobalEPKvS3_mbmm -_ZN4sycl3_V17handler28setStateExplicitKernelBundleEv _ZN4sycl3_V17handler28setArgsToAssociatedAccessorsEv +_ZN4sycl3_V17handler28setStateExplicitKernelBundleEv _ZN4sycl3_V17handler30memcpyFromHostOnlyDeviceGlobalEPvPKvbmm _ZN4sycl3_V17handler30verifyUsedKernelBundleInternalENS0_6detail11string_viewE _ZN4sycl3_V17handler32verifyDeviceHasProgressGuaranteeENS0_3ext6oneapi12experimental26forward_progress_guaranteeENS4_15execution_scopeES6_ diff --git a/sycl/test/abi/sycl_symbols_windows.dump b/sycl/test/abi/sycl_symbols_windows.dump index 7a97c440c5fa6..2e1796b9a4568 100644 --- a/sycl/test/abi/sycl_symbols_windows.dump +++ b/sycl/test/abi/sycl_symbols_windows.dump @@ -621,15 +621,17 @@ ??_Fqueue@_V1@sycl@@QEAAXXZ ?AccessTargetMask@handler@_V1@sycl@@0HB ?Clear@exception_list@_V1@sycl@@AEAAXXZ -?clearArgs@handler@_V1@sycl@@AEAAXXZ ?DirSep@OSUtil@detail@_V1@sycl@@2QEBDEB ?DisableRangeRounding@handler@_V1@sycl@@AEAA_NXZ ?GDBMethodsAnchor@SampledImageAccessorBaseHost@detail@_V1@sycl@@IEAAXXZ ?GDBMethodsAnchor@UnsampledImageAccessorBaseHost@detail@_V1@sycl@@IEAAXXZ ?GetRangeRoundingSettings@handler@_V1@sycl@@AEAAXAEA_K00@Z +?HasAssociatedAccessor@handler@_V1@sycl@@AEBA_NPEAVAccessorImplHost@detail@23@W4target@access@23@@Z ?PushBack@exception_list@_V1@sycl@@AEAAX$$QEAVexception_ptr@std@@@Z ?PushBack@exception_list@_V1@sycl@@AEAAXAEBVexception_ptr@std@@@Z ?RangeRoundingTrace@handler@_V1@sycl@@AEAA_NXZ +?SetHostTask@handler@_V1@sycl@@AEAAX$$QEAV?$function@$$A6AXVinterop_handle@_V1@sycl@@@Z@std@@@Z +?SetHostTask@handler@_V1@sycl@@AEAAX$$QEAV?$function@$$A6AXXZ@std@@@Z ?__abs_diff_impl@_V1@sycl@@YA?AV?$vec@C$00@12@V312@0@Z ?__abs_diff_impl@_V1@sycl@@YA?AV?$vec@C$01@12@V312@0@Z ?__abs_diff_impl@_V1@sycl@@YA?AV?$vec@C$02@12@V312@0@Z @@ -3667,6 +3669,8 @@ ?add@device_global_map@detail@_V1@sycl@@YAXPEBXPEBD@Z ?add@host_pipe_map@detail@_V1@sycl@@YAXPEBXPEBD@Z ?add@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@QEAA?AVnode@34567@AEBVproperty_list@67@@Z +?addAccessorReq@handler@_V1@sycl@@AEAAXV?$shared_ptr@VAccessorImplHost@detail@_V1@sycl@@@std@@@Z +?addArg@handler@_V1@sycl@@AEAAXW4kernel_param_kind_t@detail@23@PEAXHH@Z ?addCounterInit@detail@_V1@sycl@@YAXAEAVhandler@23@AEAV?$shared_ptr@Vqueue_impl@detail@_V1@sycl@@@std@@AEAV?$shared_ptr@H@6@@Z ?addGraphLeafDependencies@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEAAXVnode@34567@@Z ?addHostAccessorAndWait@detail@_V1@sycl@@YAXPEAVAccessorImplHost@123@@Z @@ -3674,12 +3678,10 @@ ?addHostUnsampledImageAccessorAndWait@detail@_V1@sycl@@YAXPEAVUnsampledImageAccessorImplHost@123@@Z ?addImpl@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEAA?AVnode@34567@AEBV?$vector@Vnode@experimental@oneapi@ext@_V1@sycl@@V?$allocator@Vnode@experimental@oneapi@ext@_V1@sycl@@@std@@@std@@@Z ?addImpl@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@IEAA?AVnode@34567@V?$function@$$A6AXAEAVhandler@_V1@sycl@@@Z@std@@AEBV?$vector@Vnode@experimental@oneapi@ext@_V1@sycl@@V?$allocator@Vnode@experimental@oneapi@ext@_V1@sycl@@@std@@@std@@@Z +?addLifetimeSharedPtrStorage@handler@_V1@sycl@@AEAAXV?$shared_ptr@$$CBX@std@@@Z ?addOrReplaceAccessorProperties@buffer_plain@detail@_V1@sycl@@IEAAXAEBVproperty_list@34@@Z ?addReduction@handler@_V1@sycl@@AEAAXAEBV?$shared_ptr@$$CBX@std@@@Z ?addStream@handler@_V1@sycl@@AEAAXAEBV?$shared_ptr@Vstream_impl@detail@_V1@sycl@@@std@@@Z -?addArg@handler@_V1@sycl@@AEAAXW4kernel_param_kind_t@detail@23@PEAXHH@Z -?addLifetimeSharedPtrStorage@handler@_V1@sycl@@AEAAXV?$shared_ptr@$$CBX@std@@@Z -?addAccessorReq@handler@_V1@sycl@@AEAAXV?$shared_ptr@VAccessorImplHost@detail@_V1@sycl@@@std@@@Z ?alignedAlloc@OSUtil@detail@_V1@sycl@@SAPEAX_K0@Z ?alignedFree@OSUtil@detail@_V1@sycl@@SAXPEAX@Z ?aligned_alloc@_V1@sycl@@YAPEAX_K0AEBVdevice@12@AEBVcontext@12@W4alloc@usm@12@AEBUcode_location@detail@12@@Z @@ -3713,10 +3715,11 @@ ?begin@kernel_bundle_plain@detail@_V1@sycl@@IEBAPEBVdevice_image_plain@234@XZ ?begin_recording@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@QEAAXAEAVqueue@67@AEBVproperty_list@67@@Z ?begin_recording@modifiable_command_graph@detail@experimental@oneapi@ext@_V1@sycl@@QEAAXAEBV?$vector@Vqueue@_V1@sycl@@V?$allocator@Vqueue@_V1@sycl@@@std@@@std@@AEBVproperty_list@67@@Z -?build_from_source@detail@experimental@oneapi@ext@_V1@sycl@@YA?AV?$kernel_bundle@$01@56@AEAV?$kernel_bundle@$02@56@AEBV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@std@@AEBV?$vector@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V?$allocator@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@@2@@std@@PEAV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@2@Z +?build_from_source@detail@experimental@oneapi@ext@_V1@sycl@@YA?AV?$kernel_bundle@$01@56@AEAV?$kernel_bundle@$02@56@AEBV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@std@@AEBV?$vector@Vstring_view@detail@_V1@sycl@@V?$allocator@Vstring_view@detail@_V1@sycl@@@std@@@std@@PEAVstring@156@2@Z ?build_impl@detail@_V1@sycl@@YA?AV?$shared_ptr@Vkernel_bundle_impl@detail@_V1@sycl@@@std@@AEBV?$kernel_bundle@$0A@@23@AEBV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@5@AEBVproperty_list@23@@Z ?cancel_fusion@fusion_wrapper@experimental@codeplay@ext@_V1@sycl@@QEAAXXZ ?category@exception@_V1@sycl@@QEBAAEBVerror_category@std@@XZ +?clearArgs@handler@_V1@sycl@@AEAAXXZ ?code@exception@_V1@sycl@@QEBAAEBVerror_code@std@@XZ ?compile_impl@detail@_V1@sycl@@YA?AV?$shared_ptr@Vkernel_bundle_impl@detail@_V1@sycl@@@std@@AEBV?$kernel_bundle@$0A@@23@AEBV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@5@AEBVproperty_list@23@@Z ?complete_fusion@fusion_wrapper@experimental@codeplay@ext@_V1@sycl@@QEAA?AVevent@56@AEBVproperty_list@56@@Z @@ -3804,6 +3807,7 @@ ?ext_oneapi_get_composite_devices@platform@_V1@sycl@@QEBA?AV?$vector@Vdevice@_V1@sycl@@V?$allocator@Vdevice@_V1@sycl@@@std@@@std@@XZ ?ext_oneapi_get_default_context@platform@_V1@sycl@@QEBA?AVcontext@23@XZ ?ext_oneapi_get_graph@queue@_V1@sycl@@QEBA?AV?$command_graph@$0A@@experimental@oneapi@ext@23@XZ +?ext_oneapi_get_kernel@kernel_bundle_plain@detail@_V1@sycl@@AEAA?AVkernel@34@Vstring_view@234@@Z ?ext_oneapi_get_kernel@kernel_bundle_plain@detail@_V1@sycl@@QEAA?AVkernel@34@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@@Z ?ext_oneapi_get_last_event@queue@_V1@sycl@@QEBA?AVevent@23@XZ ?ext_oneapi_get_state@queue@_V1@sycl@@QEBA?AW4queue_state@experimental@oneapi@ext@23@XZ @@ -3811,6 +3815,7 @@ ?ext_oneapi_graph@queue@_V1@sycl@@QEAA?AVevent@23@V?$command_graph@$00@experimental@oneapi@ext@23@AEBUcode_location@detail@23@@Z ?ext_oneapi_graph@queue@_V1@sycl@@QEAA?AVevent@23@V?$command_graph@$00@experimental@oneapi@ext@23@AEBV?$vector@Vevent@_V1@sycl@@V?$allocator@Vevent@_V1@sycl@@@std@@@std@@AEBUcode_location@detail@23@@Z ?ext_oneapi_graph@queue@_V1@sycl@@QEAA?AVevent@23@V?$command_graph@$00@experimental@oneapi@ext@23@V423@AEBUcode_location@detail@23@@Z +?ext_oneapi_has_kernel@kernel_bundle_plain@detail@_V1@sycl@@AEAA_NVstring_view@234@@Z ?ext_oneapi_has_kernel@kernel_bundle_plain@detail@_V1@sycl@@QEAA_NAEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@@Z ?ext_oneapi_memcpy2d_impl@handler@_V1@sycl@@AEAAXPEAX_KPEBX111@Z ?ext_oneapi_memset2d_impl@handler@_V1@sycl@@AEAAXPEAX_KH11@Z @@ -4048,7 +4053,6 @@ ?get_width@stream@_V1@sycl@@QEBA_KXZ ?get_work_item_buffer_size@stream@_V1@sycl@@QEBA_KXZ ?gpu_selector_v@_V1@sycl@@YAHAEBVdevice@12@@Z -?HasAssociatedAccessor@handler@_V1@sycl@@AEBA_NPEAVAccessorImplHost@detail@23@W4target@access@23@@Z ?handleRelease@buffer_plain@detail@_V1@sycl@@IEBAXXZ ?has@device@_V1@sycl@@QEBA_NW4aspect@23@@Z ?has@platform@_V1@sycl@@QEBA_NW4aspect@23@@Z @@ -4104,8 +4108,8 @@ ?make_kernel@detail@_V1@sycl@@YA?AVkernel@23@_KAEBVcontext@23@W4backend@23@@Z ?make_kernel_bundle@detail@_V1@sycl@@YA?AV?$shared_ptr@Vkernel_bundle_impl@detail@_V1@sycl@@@std@@_KAEBVcontext@23@W4bundle_state@23@W4backend@23@@Z ?make_kernel_bundle@detail@_V1@sycl@@YA?AV?$shared_ptr@Vkernel_bundle_impl@detail@_V1@sycl@@@std@@_KAEBVcontext@23@_NW4bundle_state@23@W4backend@23@@Z -?make_kernel_bundle_from_source@detail@experimental@oneapi@ext@_V1@sycl@@YA?AV?$kernel_bundle@$02@56@AEBVcontext@56@W4source_language@23456@AEBV?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V?$vector@U?$pair@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V12@@std@@V?$allocator@U?$pair@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V12@@std@@@2@@std@@@Z -?make_kernel_bundle_from_source@detail@experimental@oneapi@ext@_V1@sycl@@YA?AV?$kernel_bundle@$02@56@AEBVcontext@56@W4source_language@23456@AEBV?$vector@W4byte@std@@V?$allocator@W4byte@std@@@2@@std@@V?$vector@U?$pair@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V12@@std@@V?$allocator@U?$pair@V?$basic_string@DU?$char_traits@D@std@@V?$allocator@D@2@@std@@V12@@std@@@2@@std@@@Z +?make_kernel_bundle_from_source@detail@experimental@oneapi@ext@_V1@sycl@@YA?AV?$kernel_bundle@$02@56@AEBVcontext@56@W4source_language@23456@AEBV?$vector@W4byte@std@@V?$allocator@W4byte@std@@@2@@std@@V?$vector@U?$pair@Vstring_view@detail@_V1@sycl@@V1234@@std@@V?$allocator@U?$pair@Vstring_view@detail@_V1@sycl@@V1234@@std@@@2@@std@@@Z +?make_kernel_bundle_from_source@detail@experimental@oneapi@ext@_V1@sycl@@YA?AV?$kernel_bundle@$02@56@AEBVcontext@56@W4source_language@23456@Vstring_view@156@V?$vector@U?$pair@Vstring_view@detail@_V1@sycl@@V1234@@std@@V?$allocator@U?$pair@Vstring_view@detail@_V1@sycl@@V1234@@std@@@2@@std@@@Z ?make_platform@detail@_V1@sycl@@YA?AVplatform@23@_KW4backend@23@@Z ?make_queue@detail@_V1@sycl@@YA?AVqueue@23@_KHAEBVcontext@23@PEBVdevice@23@_NAEBVproperty_list@23@AEBV?$function@$$A6AXVexception_list@_V1@sycl@@@Z@std@@W4backend@23@@Z ?malloc@_V1@sycl@@YAPEAX_KAEBVdevice@12@AEBVcontext@12@W4alloc@usm@12@AEBUcode_location@detail@12@@Z @@ -4201,23 +4205,21 @@ ?setArgHelper@handler@_V1@sycl@@AEAAXH$$QEAVraw_kernel_arg@experimental@oneapi@ext@23@@Z ?setArgHelper@handler@_V1@sycl@@AEAAXH$$QEAVsampler@23@@Z ?setArgsHelper@handler@_V1@sycl@@AEAAXH@Z +?setArgsToAssociatedAccessors@handler@_V1@sycl@@AEAAXXZ ?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@@AEAAXW4_pi_kernel_cache_config@@@Z +?setKernelClusterLaunch@handler@_V1@sycl@@AEAAXV?$range@$02@23@H@Z ?setKernelIsCooperative@handler@_V1@sycl@@AEAAX_N@Z ?setLocalAccessorArgHelper@handler@_V1@sycl@@AEAAXHAEAVLocalAccessorBaseHost@detail@23@@Z +?setNDRangeDescriptorPadded@handler@_V1@sycl@@AEAAXV?$range@$02@23@0V?$id@$02@23@H@Z +?setNDRangeDescriptorPadded@handler@_V1@sycl@@AEAAXV?$range@$02@23@V?$id@$02@23@H@Z +?setNDRangeDescriptorPadded@handler@_V1@sycl@@AEAAXV?$range@$02@23@_NH@Z ?setNDRangeUsed@handler@_V1@sycl@@AEAAX_N@Z ?setStateExplicitKernelBundle@handler@_V1@sycl@@AEAAXXZ ?setStateSpecConstSet@handler@_V1@sycl@@AEAAXXZ -?setUserFacingNodeType@handler@_V1@sycl@@AEAAXW4node_type@experimental@oneapi@ext@23@@Z -?setNDRangeDescriptorPadded@handler@_V1@sycl@@AEAAXV?$range@$02@23@V?$id@$02@23@H@Z -?setKernelClusterLaunch@handler@_V1@sycl@@AEAAXV?$range@$02@23@H@Z -?setArgsToAssociatedAccessors@handler@_V1@sycl@@AEAAXXZ ?setType@handler@_V1@sycl@@AEAAXW4CGType@detail@23@@Z -?setNDRangeDescriptorPadded@handler@_V1@sycl@@AEAAXV?$range@$02@23@_NH@Z -?SetHostTask@handler@_V1@sycl@@AEAAX$$QEAV?$function@$$A6AXVinterop_handle@_V1@sycl@@@Z@std@@@Z -?SetHostTask@handler@_V1@sycl@@AEAAX$$QEAV?$function@$$A6AXXZ@std@@@Z -?setNDRangeDescriptorPadded@handler@_V1@sycl@@AEAAXV?$range@$02@23@0V?$id@$02@23@H@Z +?setUserFacingNodeType@handler@_V1@sycl@@AEAAXW4node_type@experimental@oneapi@ext@23@@Z ?set_access_mode@experimental@oneapi@ext@_V1@sycl@@YAXPEBX_KW4address_access_mode@12345@AEBVcontext@45@@Z ?set_arg@handler@_V1@sycl@@QEAAXH$$QEAVraw_kernel_arg@experimental@oneapi@ext@23@@Z ?set_final_data_internal@buffer_plain@detail@_V1@sycl@@IEAAXAEBV?$function@$$A6AXAEBV?$function@$$A6AXPEAX@Z@std@@@Z@std@@@Z