Skip to content
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
27 commits
Select commit Hold shift + click to select a range
a06f21e
[SYCL] ABI-neutralize kernel-bundle
bso-intel Mar 30, 2024
73fe121
no export
bso-intel Apr 1, 2024
0b7554f
Merge remote-tracking branch 'upstream/sycl' into has-kernel
bso-intel Apr 15, 2024
23f6729
inline
bso-intel Apr 18, 2024
c527fe4
string_view
bso-intel Apr 18, 2024
56554a9
export
bso-intel Apr 18, 2024
6b85d7b
dump symbol
bso-intel Apr 18, 2024
0a52fe1
error
bso-intel Apr 18, 2024
0771188
dump symbol
bso-intel Apr 18, 2024
2882678
string return
bso-intel Apr 19, 2024
4239f2a
Merge branch 'has-kernel' of https://github.com/bso-intel/llvm into h…
bso-intel Apr 19, 2024
d8d3ac5
dump
bso-intel Apr 19, 2024
3c628e6
Merge remote-tracking branch 'upstream/sycl' into has-kernel
bso-intel Jul 11, 2024
4839588
fix abi symbols
bso-intel Jul 11, 2024
1236db7
fixed byte vector
bso-intel Jul 11, 2024
c8e914d
fixed byte vector
bso-intel Jul 11, 2024
fe03492
win symbol
bso-intel Jul 15, 2024
42f282b
Merge remote-tracking branch 'upstream/sycl' into has-kernel
bso-intel Jul 15, 2024
5c2eeb2
removed exclusion
bso-intel Jul 15, 2024
ee512f8
Merge branch 'sycl' into has-kernel
bso-intel Jul 18, 2024
46bc3f6
Update sycl/include/sycl/kernel_bundle.hpp
bso-intel Jul 18, 2024
35629fc
Update sycl/include/sycl/kernel_bundle.hpp
bso-intel Jul 18, 2024
26d1ab1
Update sycl/include/sycl/kernel_bundle.hpp
bso-intel Jul 18, 2024
8b1d815
Merge remote-tracking branch 'upstream/sycl' into has-kernel
bso-intel Jul 18, 2024
8ab94d5
abi symbols removed
bso-intel Jul 18, 2024
b8497e1
abi symbol
bso-intel Jul 18, 2024
468236f
win symbols
bso-intel Jul 18, 2024
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
80 changes: 73 additions & 7 deletions sycl/include/sycl/kernel_bundle.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -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
Expand All @@ -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
Expand Down Expand Up @@ -897,26 +905,84 @@ __SYCL_EXPORT bool is_source_kernel_bundle_supported(backend BE,
source_language Language);

__SYCL_EXPORT kernel_bundle<bundle_state::ext_oneapi_source>
make_kernel_bundle_from_source(
const context &SyclContext, source_language Language,
sycl::detail::string_view Source,
std::vector<std::pair<sycl::detail::string_view, sycl::detail::string_view>>
IncludePairsVec);

inline kernel_bundle<bundle_state::ext_oneapi_source>
make_kernel_bundle_from_source(
const context &SyclContext, source_language Language,
const std::string &Source,
std::vector<std::pair<std::string, std::string>> IncludePairsVec);
std::vector<std::pair<std::string, std::string>> IncludePairsVec) {
size_t n = IncludePairsVec.size();
std::vector<std::pair<sycl::detail::string_view, sycl::detail::string_view>>
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<bundle_state::ext_oneapi_source>
make_kernel_bundle_from_source(
const context &SyclContext, source_language Language,
const std::vector<std::byte> &Bytes,
std::vector<std::pair<std::string, std::string>> IncludePairsVec);
std::vector<std::pair<sycl::detail::string_view, sycl::detail::string_view>>
IncludePairsVec);

inline kernel_bundle<bundle_state::ext_oneapi_source>
make_kernel_bundle_from_source(
const context &SyclContext, source_language Language,
const std::vector<std::byte> &Bytes,
std::vector<std::pair<std::string, std::string>> IncludePairsVec) {
size_t n = IncludePairsVec.size();
std::vector<std::pair<sycl::detail::string_view, sycl::detail::string_view>>
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<bundle_state::executable>
__SYCL_EXPORT kernel_bundle<bundle_state::executable> build_from_source(
kernel_bundle<bundle_state::ext_oneapi_source> &SourceKB,
const std::vector<device> &Devices,
const std::vector<sycl::detail::string_view> &BuildOptions,
sycl::detail::string *LogPtr,
const std::vector<sycl::detail::string_view> &RegisteredKernelNames);

inline kernel_bundle<bundle_state::executable>
build_from_source(kernel_bundle<bundle_state::ext_oneapi_source> &SourceKB,
const std::vector<device> &Devices,
const std::vector<std::string> &BuildOptions,
std::string *LogPtr,
const std::vector<std::string> &RegisteredKernelNames);

const std::vector<std::string> &RegisteredKernelNames) {
std::vector<sycl::detail::string_view> Options;
for (const std::string &opt : BuildOptions)
Options.push_back(sycl::detail::string_view{opt});

std::vector<sycl::detail::string_view> 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

/////////////////////////
Expand Down
59 changes: 42 additions & 17 deletions sycl/source/kernel_bundle.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -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);
}

//////////////////////////////////
Expand Down Expand Up @@ -391,14 +391,24 @@ bool is_source_kernel_bundle_supported(backend BE, source_language Language) {
/////////////////////////

using include_pairs_t = std::vector<std::pair<std::string, std::string>>;

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<sycl::detail::string_view, sycl::detail::string_view>>;

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),
Expand All @@ -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<std::byte> &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))
Expand All @@ -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<device> &Devices,
const std::vector<std::string> &BuildOptions,
std::string *LogPtr,
const std::vector<std::string> &RegisteredKernelNames) {
exe_kb build_from_source(
source_kb &SourceKB, const std::vector<device> &Devices,
const std::vector<sycl::detail::string_view> &BuildOptions,
sycl::detail::string *LogView,
const std::vector<sycl::detail::string_view> &RegisteredKernelNames) {
std::vector<std::string> Options;
for (const sycl::detail::string_view option : BuildOptions)
Options.push_back(option.data());

std::vector<std::string> 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<device> UniqueDevices =
sycl::detail::removeDuplicateDevices(Devices);
std::shared_ptr<kernel_bundle_impl> sourceImpl = getSyclObjImpl(SourceKB);
std::shared_ptr<kernel_bundle_impl> KBImpl = sourceImpl->build_from_source(
UniqueDevices, BuildOptions, LogPtr, RegisteredKernelNames);
return sycl::detail::createSyclObjFromImpl<exe_kb>(KBImpl);
UniqueDevices, Options, LogPtr, KernelNames);
auto result = sycl::detail::createSyclObjFromImpl<exe_kb>(KBImpl);
if (LogView)
*LogView = Log;
return result;
}

} // namespace detail
Expand Down
5 changes: 0 additions & 5 deletions sycl/test/abi/sycl_abi_neutrality_test.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -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
Expand Down
18 changes: 9 additions & 9 deletions sycl/test/abi/sycl_symbols_linux.dump
Original file line number Diff line number Diff line change
Expand Up @@ -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
Expand All @@ -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
Expand Down Expand Up @@ -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
Expand Down Expand Up @@ -3501,24 +3501,24 @@ _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
_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_
Expand Down
Loading