35 #include <type_traits>
41 inline namespace _V1 {
43 template <backend Backend>
class backend_traits;
44 template <backend Backend, bundle_state State>
45 auto get_native(
const kernel_bundle<State> &Obj)
46 -> backend_return_t<Backend, kernel_bundle<State>>;
63 const char *get_name() const
noexcept;
65 bool operator==(const
kernel_id &RHS)
const {
return impl == RHS.impl; }
72 kernel_id(
const std::shared_ptr<detail::kernel_id_impl> &Impl)
73 : impl(
std::move(Impl)) {}
75 std::shared_ptr<detail::kernel_id_impl> impl;
85 class device_image_impl;
93 : impl(
std::move(Impl)) {}
96 return impl == RHS.
impl;
100 return !(*
this == RHS);
121 template <sycl::bundle_state State>
130 return device_image_plain::has_kernel(KernelID);
136 return device_image_plain::has_kernel(KernelID, Dev);
158 : impl(
std::move(Impl)) {}
161 return impl == RHS.
impl;
165 return !(*
this == RHS);
182 bool contains_specialization_constants() const
noexcept;
184 bool native_specialization_constant() const
noexcept;
186 bool ext_oneapi_has_kernel(const
std::
string &name);
188 kernel ext_oneapi_get_kernel(const
std::
string &name);
201 bool has_specialization_constant_impl(const
char *SpecName) const
noexcept;
203 void set_specialization_constant_impl(const
char *SpecName,
void *Value,
206 void get_specialization_constant_impl(const
char *SpecName,
211 bool is_specialization_constant_set(const
char *SpecName) const
noexcept;
233 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
257 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
267 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
276 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
278 return has_kernel(get_kernel_id<KernelName>());
285 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
287 return has_kernel(get_kernel_id<KernelName>(), Dev);
293 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
302 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
311 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
319 typename = std::enable_if_t<_State == bundle_state::executable>>
326 template <
typename KernelName,
bundle_state _State = State,
327 typename = std::enable_if_t<_State == bundle_state::executable>>
336 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
338 const char *SpecSymName = detail::get_spec_constant_symbolic_ID<SpecName>();
339 return has_specialization_constant_impl(SpecSymName);
346 typename = std::enable_if_t<_State == bundle_state::input>>
348 typename std::remove_reference_t<decltype(SpecName)>::
value_type Value) {
349 const char *SpecSymName = detail::get_spec_constant_symbolic_ID<SpecName>();
350 set_specialization_constant_impl(SpecSymName, &Value,
351 sizeof(decltype(Value)));
358 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
359 typename std::remove_reference_t<decltype(SpecName)>
::value_type
362 typename std::remove_reference_t<decltype(SpecName)>
::value_type;
364 const char *SpecSymName = detail::get_spec_constant_symbolic_ID<SpecName>();
365 SCType Res{SpecName.getDefaultValue()};
366 if (!is_specialization_constant_set(SpecSymName))
369 std::array<char,
sizeof(SCType)> RetValue;
370 get_specialization_constant_impl(SpecSymName, RetValue.data());
371 std::memcpy(&Res, RetValue.data(),
sizeof(SCType));
379 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
388 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
398 typename = std::enable_if_t<_State == bundle_state::executable>>
408 typename = std::enable_if_t<_State == bundle_state::executable>>
415 : kernel_bundle_plain(
std::move(Impl)) {}
423 template <backend Backend, bundle_state StateB>
427 template <backend Backend>
432 ReturnValue.reserve(std::distance(begin(), end()));
435 ReturnValue.push_back(
443 template <bundle_state State>
460 using KI = sycl::detail::KernelInfo<KernelName>;
479 __SYCL_EXPORT
const std::vector<device>
488 template <bundle_state State>
490 const std::vector<device> &Devs) {
496 return detail::createSyclObjFromImpl<kernel_bundle<State>>(Impl);
499 template <bundle_state State>
501 return get_kernel_bundle<State>(Ctx, Ctx.
get_devices());
510 const std::vector<kernel_id> &KernelIDs,
522 template <bundle_state State>
525 const std::vector<kernel_id> &KernelIDs) {
530 return detail::createSyclObjFromImpl<kernel_bundle<State>>(Impl);
533 template <bundle_state State>
536 return get_kernel_bundle<State>(Ctx, Ctx.
get_devices(), KernelIDs);
539 template <
typename KernelName, bundle_state State>
541 return get_kernel_bundle<State>(Ctx, Ctx.
get_devices(),
542 {get_kernel_id<KernelName>()});
545 template <
typename KernelName, bundle_state State>
547 const std::vector<device> &Devs) {
548 return get_kernel_bundle<State>(Ctx, Devs, {get_kernel_id<KernelName>()});
567 const std::vector<device> &Devs);
571 template <bundle_state State>
575 return detail::createSyclObjFromImpl<sycl::kernel_bundle<State>>(Impl);
581 template <bundle_state State,
typename SelectorT>
583 const std::vector<device> &Devs,
584 SelectorT Selector) {
594 Ctx, UniqueDevices, State, SelectorWrapper);
596 return detail::createSyclObjFromImpl<sycl::kernel_bundle<State>>(Impl);
599 template <bundle_state State,
typename SelectorT>
601 return get_kernel_bundle<State>(Ctx, Ctx.
get_devices(), Selector);
611 const std::vector<device> &Devs,
616 const std::vector<kernel_id> &kernelIds,
630 template <bundle_state State>
635 template <bundle_state State>
637 const std::vector<kernel_id> &KernelIDs) {
642 return has_kernel_bundle<State>(Ctx, Ctx.
get_devices());
645 template <bundle_state State>
647 const std::vector<kernel_id> &KernelIDs) {
648 return has_kernel_bundle<State>(Ctx, Ctx.
get_devices(), KernelIDs);
651 template <
typename KernelName, bundle_state State>
653 return has_kernel_bundle<State>(Ctx, {get_kernel_id<KernelName>()});
656 template <
typename KernelName, bundle_state State>
658 return has_kernel_bundle<State>(Ctx, Devs, {get_kernel_id<KernelName>()});
667 __SYCL_EXPORT
bool is_compatible(
const std::vector<kernel_id> &KernelIDs,
680 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
681 join_impl(
const std::vector<detail::KernelBundleImplPtr> &Bundles,
687 template <sycl::bundle_state State>
691 std::vector<detail::KernelBundleImplPtr> KernelBundleImpls;
692 KernelBundleImpls.reserve(Bundles.size());
696 std::shared_ptr<detail::kernel_bundle_impl> Impl =
698 return detail::createSyclObjFromImpl<kernel_bundle<State>>(Impl);
707 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
708 compile_impl(
const kernel_bundle<bundle_state::input> &InputBundle,
709 const std::vector<device> &Devs,
const property_list &PropList);
716 inline kernel_bundle<bundle_state::object>
718 const std::vector<device> &Devs,
const property_list &PropList = {}) {
724 kernel_bundle<sycl::bundle_state::object>>(Impl);
727 inline kernel_bundle<bundle_state::object>
739 const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles);
741 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
742 link_impl(
const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles,
743 const std::vector<device> &Devs,
const property_list &PropList);
751 inline kernel_bundle<bundle_state::executable>
753 const std::vector<device> &Devs,
const property_list &PropList = {}) {
759 kernel_bundle<sycl::bundle_state::executable>>(Impl);
762 inline kernel_bundle<bundle_state::executable>
765 return link(std::vector<kernel_bundle<bundle_state::object>>{ObjectBundle},
769 inline kernel_bundle<bundle_state::executable>
772 std::vector<sycl::device> IntersectDevices =
774 return link(ObjectBundles, IntersectDevices, PropList);
777 inline kernel_bundle<bundle_state::executable>
779 const std::vector<device> &Devs,
const property_list &PropList = {}) {
780 return link(std::vector<kernel_bundle<bundle_state::object>>{ObjectBundle},
789 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
790 build_impl(
const kernel_bundle<bundle_state::input> &InputBundle,
791 const std::vector<device> &Devs,
const property_list &PropList);
798 inline kernel_bundle<bundle_state::executable>
800 const std::vector<device> &Devs,
const property_list &PropList = {}) {
806 kernel_bundle<sycl::bundle_state::executable>>(Impl);
809 inline kernel_bundle<bundle_state::executable>
812 return build(InputBundle, InputBundle.
get_devices(), PropList);
815 namespace ext::oneapi::experimental {
860 const std::string &Source);
862 #if (!defined(_HAS_STD_BYTE) || _HAS_STD_BYTE != 0)
866 const std::vector<std::byte> &Bytes);
876 const std::vector<device> &Devices,
878 std::string *LogPtr);
883 typename = std::enable_if_t<
884 is_property_list_v<PropertyListT> &&
887 PropertyListT>::value>>
891 const std::vector<device> &Devices, PropertyListT props = {}) {
892 std::vector<std::string> BuildOptionsVec;
893 std::string *LogPtr =
nullptr;
894 if constexpr (props.template has_property<build_options>()) {
895 BuildOptionsVec = props.template get_property<build_options>().opts;
897 if constexpr (props.template has_property<save_log>()) {
898 LogPtr = props.template get_property<save_log>().log;
904 typename = std::enable_if_t<
905 is_property_list_v<PropertyListT> &&
906 detail::all_props_are_keys_of<
908 PropertyListT>::value>>
911 PropertyListT props = {}) {
912 return build<PropertyListT>(SourceKB, SourceKB.
get_devices(), props);
921 template <>
struct hash<
sycl::kernel_id> {
923 return hash<std::shared_ptr<sycl::detail::kernel_id_impl>>()(
928 template <sycl::bundle_state State>
struct hash<
sycl::device_image<State>> {
930 return hash<std::shared_ptr<sycl::detail::device_image_impl>>()(
937 return hash<std::shared_ptr<sycl::detail::kernel_bundle_impl>>()(
The context class represents a SYCL context on which kernel functions may be executed.
std::vector< device > get_devices() const
Gets devices associated with this SYCL context.
bool operator==(const device_image_plain &RHS) const
device_image_plain(const detail::DeviceImageImplPtr &Impl)
bool operator!=(const device_image_plain &RHS) const
detail::DeviceImageImplPtr impl
pi_native_handle getNative() const
bool contains_specialization_constants() const noexcept
std::vector< kernel_id > get_kernel_ids() const
bool operator!=(const kernel_bundle_plain &RHS) const
bool ext_oneapi_has_kernel(const std::string &name)
const device_image_plain * end() const
bool native_specialization_constant() const noexcept
context get_context() const noexcept
bool empty() const noexcept
detail::KernelBundleImplPtr impl
std::vector< device > get_devices() const noexcept
kernel_bundle_plain(const detail::KernelBundleImplPtr &Impl)
backend get_backend() const noexcept
bool operator==(const kernel_bundle_plain &RHS) const
const device_image_plain * begin() const
bool has_kernel(const kernel_id &KernelID) const noexcept
kernel ext_oneapi_get_kernel(const std::string &name)
kernel get_kernel(const kernel_id &KernelID) const
Objects of the class represents an instance of an image in a specific state.
bool has_kernel(const kernel_id &KernelID, const device &Dev) const noexcept
bool has_kernel(const kernel_id &KernelID) const noexcept
The SYCL device class encapsulates a single SYCL device on which kernels may be executed.
The kernel_bundle class represents collection of device images in a particular state.
bool ext_oneapi_has_kernel(const std::string &name)
std::vector< kernel_id > get_kernel_ids() const
bool contains_specialization_constants() const noexcept
bool has_kernel(const kernel_id &KernelID) const noexcept
bool has_specialization_constant() const noexcept
device_image_iterator begin() const
kernel get_kernel() const
device_image_iterator end() const
kernel get_kernel(const kernel_id &KernelID) const
bool has_kernel(const device &Dev) const noexcept
kernel ext_oneapi_get_kernel(const std::string &name)
backend get_backend() const noexcept
friend auto get_native(const kernel_bundle< StateB > &Obj) -> backend_return_t< Backend, kernel_bundle< StateB >>
std::vector< device > get_devices() const noexcept
bool has_kernel(const kernel_id &KernelID, const device &Dev) const noexcept
context get_context() const noexcept
bool has_kernel() const noexcept
bool empty() const noexcept
bool native_specialization_constant() const noexcept
std::remove_reference_t< decltype(SpecName)>::value_type get_specialization_constant() const
void set_specialization_constant(typename std::remove_reference_t< decltype(SpecName)>::value_type Value)
Sets the value of the specialization constant whose address is SpecName for this bundle.
Objects of the class identify kernel is some kernel_bundle related APIs.
bool operator!=(const kernel_id &RHS) const
Provides an abstraction of a SYCL kernel.
Objects of the property_list class are containers for the SYCL properties.
PiProgram cast(cl_program)=delete
std::vector< sycl::device > find_device_intersection(const std::vector< kernel_bundle< bundle_state::object >> &ObjectBundles)
std::shared_ptr< detail::kernel_bundle_impl > build_impl(const kernel_bundle< bundle_state::input > &InputBundle, const std::vector< device > &Devs, const property_list &PropList)
std::shared_ptr< device_image_impl > DeviceImageImplPtr
std::function< bool(const detail::DeviceImageImplPtr &DevImgImpl)> DevImgSelectorImpl
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
const std::vector< device > removeDuplicateDevices(const std::vector< device > &Devs)
detail::KernelBundleImplPtr get_kernel_bundle_impl(const context &Ctx, const std::vector< device > &Devs, bundle_state State)
T createSyclObjFromImpl(decltype(T::impl) ImplObj)
std::shared_ptr< detail::kernel_bundle_impl > compile_impl(const kernel_bundle< bundle_state::input > &InputBundle, const std::vector< device > &Devs, const property_list &PropList)
std::shared_ptr< detail::kernel_bundle_impl > join_impl(const std::vector< detail::KernelBundleImplPtr > &Bundles, bundle_state State)
kernel_id get_kernel_id_impl(string_view KernelName)
std::shared_ptr< detail::kernel_bundle_impl > KernelBundleImplPtr
kernel_bundle< State > get_empty_interop_kernel_bundle(const context &Ctx)
make_kernel may need an empty interop kernel bundle.
bool has_kernel_bundle_impl(const context &Ctx, const std::vector< device > &Devs, bundle_state State)
detail::KernelBundleImplPtr get_empty_interop_kernel_bundle_impl(const context &Ctx, const std::vector< device > &Devs)
std::shared_ptr< detail::kernel_bundle_impl > link_impl(const std::vector< kernel_bundle< bundle_state::object >> &ObjectBundles, const std::vector< device > &Devs, const property_list &PropList)
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)
kernel_bundle< bundle_state::ext_oneapi_source > create_kernel_bundle_from_source(const context &SyclContext, source_language Language, const std::vector< std::byte > &Bytes)
bool is_source_kernel_bundle_supported(backend BE, source_language Language)
sycl::detail::kernel_bundle_impl kernel_bundle_impl
decltype(properties{}) empty_properties_t
kernel_bundle(kernel_bundle< State > &&) -> kernel_bundle< State >
sycl::kernel_bundle< State > join(const std::vector< sycl::kernel_bundle< State >> &Bundles)
kernel_bundle< bundle_state::executable > build(const kernel_bundle< bundle_state::input > &InputBundle, const std::vector< device > &Devs, const property_list &PropList={})
kernel_id get_kernel_id()
kernel_bundle< bundle_state::object > compile(const kernel_bundle< bundle_state::input > &InputBundle, const std::vector< device > &Devs, const property_list &PropList={})
std::vector< kernel_id > get_kernel_ids()
bool is_compatible(const std::vector< kernel_id > &KernelIDs, const device &Dev)
bool has_kernel_bundle(const context &Ctx, const std::vector< device > &Devs)
kernel_bundle< State > get_kernel_bundle(const context &Ctx, const std::vector< device > &Devs)
A kernel bundle in state State which contains all of the kernels in the application which are compati...
ESIMD_NODEBUG ESIMD_INLINE sycl::ext::intel::esimd::simd< float, SZ > log(sycl::ext::intel::esimd::simd< float, SZ > x) __NOEXC
auto get_native(const SyclObjectT &Obj) -> backend_return_t< BackendName, SyclObjectT >
typename backend_traits< Backend >::template return_type< SyclType > backend_return_t
kernel_bundle< bundle_state::executable > link(const std::vector< kernel_bundle< bundle_state::object >> &ObjectBundles, const std::vector< device > &Devs, const property_list &PropList={})
uintptr_t pi_native_handle
C++ wrapper of extern "C" PI interfaces.
_Abi const simd< _Tp, _Abi > & noexcept
size_t operator()(const sycl::device_image< State > &DeviceImage) const
size_t operator()(const sycl::kernel_bundle< State > &KernelBundle) const
size_t operator()(const sycl::kernel_id &KernelID) const
build_options(const std::string &optsArg)
std::vector< std::string > opts
build_options(const std::vector< std::string > &optsArg)
save_log(std::string *logArg)