18 #ifdef __INTEL_PREVIEW_BREAKING_CHANGES
37 #include <type_traits>
43 inline namespace _V1 {
45 template <backend Backend>
class backend_traits;
46 template <backend Backend, bundle_state State>
47 auto get_native(
const kernel_bundle<State> &Obj)
48 -> backend_return_t<Backend, kernel_bundle<State>>;
65 const char *get_name() const
noexcept;
67 bool operator==(const
kernel_id &RHS)
const {
return impl == RHS.impl; }
74 kernel_id(
const std::shared_ptr<detail::kernel_id_impl> &Impl)
75 : impl(
std::move(Impl)) {}
77 std::shared_ptr<detail::kernel_id_impl> impl;
87 class device_image_impl;
95 : impl(
std::move(Impl)) {}
98 return impl == RHS.
impl;
102 return !(*
this == RHS);
123 template <sycl::bundle_state State>
132 return device_image_plain::has_kernel(KernelID);
138 return device_image_plain::has_kernel(KernelID, Dev);
160 : impl(
std::move(Impl)) {}
163 return impl == RHS.
impl;
167 return !(*
this == RHS);
184 bool contains_specialization_constants() const
noexcept;
186 bool native_specialization_constant() const
noexcept;
188 bool ext_oneapi_has_kernel(const
std::
string &name);
190 kernel ext_oneapi_get_kernel(const
std::
string &name);
203 bool has_specialization_constant_impl(const
char *SpecName) const
noexcept;
205 void set_specialization_constant_impl(const
char *SpecName,
void *Value,
208 void get_specialization_constant_impl(const
char *SpecName,
213 bool is_specialization_constant_set(const
char *SpecName) const
noexcept;
235 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
259 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
269 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
278 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
280 return has_kernel(get_kernel_id<KernelName>());
287 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
289 return has_kernel(get_kernel_id<KernelName>(), Dev);
295 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
304 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
313 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
321 typename = std::enable_if_t<_State == bundle_state::executable>>
328 template <
typename KernelName,
bundle_state _State = State,
329 typename = std::enable_if_t<_State == bundle_state::executable>>
338 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
340 const char *SpecSymName = detail::get_spec_constant_symbolic_ID<SpecName>();
341 return has_specialization_constant_impl(SpecSymName);
348 typename = std::enable_if_t<_State == bundle_state::input>>
350 typename std::remove_reference_t<decltype(SpecName)>::
value_type Value) {
351 const char *SpecSymName = detail::get_spec_constant_symbolic_ID<SpecName>();
352 set_specialization_constant_impl(SpecSymName, &Value,
353 sizeof(decltype(Value)));
360 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
361 typename std::remove_reference_t<decltype(SpecName)>
::value_type
364 typename std::remove_reference_t<decltype(SpecName)>
::value_type;
366 const char *SpecSymName = detail::get_spec_constant_symbolic_ID<SpecName>();
367 SCType Res{SpecName.getDefaultValue()};
368 if (!is_specialization_constant_set(SpecSymName))
371 std::array<char,
sizeof(SCType)> RetValue;
372 get_specialization_constant_impl(SpecSymName, RetValue.data());
373 std::memcpy(&Res, RetValue.data(),
sizeof(SCType));
381 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
390 typename = std::enable_if_t<_State != bundle_state::ext_oneapi_source>>
400 typename = std::enable_if_t<_State == bundle_state::executable>>
410 typename = std::enable_if_t<_State == bundle_state::executable>>
417 : kernel_bundle_plain(
std::move(Impl)) {}
425 template <backend Backend, bundle_state StateB>
429 template <backend Backend>
437 ReturnValue.push_back(
445 template <bundle_state State>
453 #ifndef __INTEL_PREVIEW_BREAKING_CHANGES
465 using KI = sycl::detail::KernelInfo<KernelName>;
466 #ifdef __INTEL_PREVIEW_BREAKING_CHANGES
488 __SYCL_EXPORT
const std::vector<device>
497 template <bundle_state State>
499 const std::vector<device> &Devs) {
505 return detail::createSyclObjFromImpl<kernel_bundle<State>>(Impl);
508 template <bundle_state State>
510 return get_kernel_bundle<State>(Ctx, Ctx.
get_devices());
519 const std::vector<kernel_id> &KernelIDs,
531 template <bundle_state State>
534 const std::vector<kernel_id> &KernelIDs) {
539 return detail::createSyclObjFromImpl<kernel_bundle<State>>(Impl);
542 template <bundle_state State>
545 return get_kernel_bundle<State>(Ctx, Ctx.
get_devices(), KernelIDs);
548 template <
typename KernelName, bundle_state State>
550 return get_kernel_bundle<State>(Ctx, Ctx.
get_devices(),
551 {get_kernel_id<KernelName>()});
554 template <
typename KernelName, bundle_state State>
556 const std::vector<device> &Devs) {
557 return get_kernel_bundle<State>(Ctx, Devs, {get_kernel_id<KernelName>()});
576 const std::vector<device> &Devs);
580 template <bundle_state State>
584 return detail::createSyclObjFromImpl<sycl::kernel_bundle<State>>(Impl);
590 template <bundle_state State,
typename SelectorT>
592 const std::vector<device> &Devs,
593 SelectorT Selector) {
603 Ctx, UniqueDevices, State, SelectorWrapper);
605 return detail::createSyclObjFromImpl<sycl::kernel_bundle<State>>(Impl);
608 template <bundle_state State,
typename SelectorT>
610 return get_kernel_bundle<State>(Ctx, Ctx.
get_devices(), Selector);
620 const std::vector<device> &Devs,
625 const std::vector<kernel_id> &kernelIds,
639 template <bundle_state State>
644 template <bundle_state State>
646 const std::vector<kernel_id> &KernelIDs) {
651 return has_kernel_bundle<State>(Ctx, Ctx.
get_devices());
654 template <bundle_state State>
656 const std::vector<kernel_id> &KernelIDs) {
657 return has_kernel_bundle<State>(Ctx, Ctx.
get_devices(), KernelIDs);
660 template <
typename KernelName, bundle_state State>
662 return has_kernel_bundle<State>(Ctx, {get_kernel_id<KernelName>()});
665 template <
typename KernelName, bundle_state State>
667 return has_kernel_bundle<State>(Ctx, Devs, {get_kernel_id<KernelName>()});
676 __SYCL_EXPORT
bool is_compatible(
const std::vector<kernel_id> &KernelIDs,
689 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
690 join_impl(
const std::vector<detail::KernelBundleImplPtr> &Bundles,
696 template <sycl::bundle_state State>
700 std::vector<detail::KernelBundleImplPtr> KernelBundleImpls;
701 KernelBundleImpls.reserve(Bundles.size());
705 std::shared_ptr<detail::kernel_bundle_impl> Impl =
707 return detail::createSyclObjFromImpl<kernel_bundle<State>>(Impl);
716 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
717 compile_impl(
const kernel_bundle<bundle_state::input> &InputBundle,
718 const std::vector<device> &Devs,
const property_list &PropList);
725 inline kernel_bundle<bundle_state::object>
727 const std::vector<device> &Devs,
const property_list &PropList = {}) {
733 kernel_bundle<sycl::bundle_state::object>>(Impl);
736 inline kernel_bundle<bundle_state::object>
748 const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles);
750 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
751 link_impl(
const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles,
752 const std::vector<device> &Devs,
const property_list &PropList);
760 inline kernel_bundle<bundle_state::executable>
762 const std::vector<device> &Devs,
const property_list &PropList = {}) {
768 kernel_bundle<sycl::bundle_state::executable>>(Impl);
771 inline kernel_bundle<bundle_state::executable>
774 return link(std::vector<kernel_bundle<bundle_state::object>>{ObjectBundle},
778 inline kernel_bundle<bundle_state::executable>
781 std::vector<sycl::device> IntersectDevices =
783 return link(ObjectBundles, IntersectDevices, PropList);
786 inline kernel_bundle<bundle_state::executable>
788 const std::vector<device> &Devs,
const property_list &PropList = {}) {
789 return link(std::vector<kernel_bundle<bundle_state::object>>{ObjectBundle},
798 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
799 build_impl(
const kernel_bundle<bundle_state::input> &InputBundle,
800 const std::vector<device> &Devs,
const property_list &PropList);
807 inline kernel_bundle<bundle_state::executable>
809 const std::vector<device> &Devs,
const property_list &PropList = {}) {
815 kernel_bundle<sycl::bundle_state::executable>>(Impl);
818 inline kernel_bundle<bundle_state::executable>
821 return build(InputBundle, InputBundle.
get_devices(), PropList);
824 namespace ext::oneapi::experimental {
871 #if (!defined(_HAS_STD_BYTE) || _HAS_STD_BYTE != 0)
875 const std::vector<std::byte> &Bytes);
885 const std::vector<device> &Devices,
892 typename = std::enable_if_t<
893 is_property_list_v<PropertyListT> &&
896 PropertyListT>::value>>
900 const std::vector<device> &Devices, PropertyListT props = {}) {
901 std::vector<std::string> BuildOptionsVec;
903 if constexpr (props.template has_property<build_options>()) {
904 BuildOptionsVec = props.template get_property<build_options>().opts;
906 if constexpr (props.template has_property<save_log>()) {
907 LogPtr = props.template get_property<save_log>().log;
913 typename = std::enable_if_t<
914 is_property_list_v<PropertyListT> &&
915 detail::all_props_are_keys_of<
917 PropertyListT>::value>>
920 PropertyListT props = {}) {
921 return build<PropertyListT>(SourceKB, SourceKB.
get_devices(), props);
930 template <>
struct hash<
sycl::kernel_id> {
932 return hash<std::shared_ptr<sycl::detail::kernel_id_impl>>()(
937 template <sycl::bundle_state State>
struct hash<
sycl::device_image<State>> {
939 return hash<std::shared_ptr<sycl::detail::device_image_impl>>()(
946 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={})
float distance(T p0, T p1)
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)