16 inline namespace _V1 {
22 : MKernel(
Kernel), MContext(Context),
25 MCreatedFromSource(true), MKernelBundleImpl(
std::move(KernelBundleImpl)),
26 MIsInterop(true), MKernelArgMaskPtr{ArgMask} {
27 ur_context_handle_t UrContext =
nullptr;
29 getPlugin()->call(urKernelGetInfo, MKernel, UR_KERNEL_INFO_CONTEXT,
30 sizeof(UrContext), &UrContext,
nullptr);
31 if (Context->getHandleRef() != UrContext)
34 "Input context must be the same as the context of cl_kernel");
39 if (Context->getPlatformImpl()->supports_usm()) {
40 bool EnableAccess =
true;
41 getPlugin()->call(urKernelSetExecInfo, MKernel,
42 UR_KERNEL_EXEC_INFO_USM_INDIRECT_ACCESS,
43 sizeof(ur_bool_t),
nullptr, &EnableAccess);
51 ur_program_handle_t Program, std::mutex *CacheMutex)
52 : MKernel(
Kernel), MContext(
std::move(ContextImpl)), MProgram(Program),
53 MCreatedFromSource(false), MDeviceImageImpl(
std::move(DeviceImageImpl)),
54 MKernelBundleImpl(
std::move(KernelBundleImpl)),
55 MKernelArgMaskPtr{ArgMask}, MCacheMutex{CacheMutex} {
56 MIsInterop = MKernelBundleImpl->isInterop();
62 getPlugin()->call(urKernelRelease, MKernel);
63 }
catch (std::exception &e) {
81 return MCreatedFromSource;
84 bool kernel_impl::isBuiltInKernel(
const device &Device)
const {
85 auto BuiltInKernels = Device.
get_info<info::device::built_in_kernel_ids>();
86 if (BuiltInKernels.empty())
88 std::string KernelName = get_info<info::kernel::function_name>();
90 BuiltInKernels.begin(), BuiltInKernels.end(),
91 [&KernelName](
kernel_id &Id) { return Id.get_name() == KernelName; }));
94 void kernel_impl::checkIfValidForNumArgsInfoQuery()
const {
95 if (MKernelBundleImpl->isInterop())
97 auto Devices = MKernelBundleImpl->get_devices();
99 [
this](
device &Device) { return isBuiltInKernel(Device); }))
104 "info::kernel::num_args descriptor may only be used to query a kernel "
105 "that resides in a kernel bundle constructed using a backend specific"
106 "interoperability function or to query a device built-in kernel");
110 typename info::platform::version::return_type
111 kernel_impl::get_backend_info<info::platform::version>()
const {
114 "the info::platform::version info descriptor can "
115 "only be queried with an OpenCL backend");
117 auto Devices = MKernelBundleImpl->get_devices();
118 return Devices[0].get_platform().get_info<info::platform::version>();
122 std::vector<device> &Devices);
125 typename info::device::version::return_type
126 kernel_impl::get_backend_info<info::device::version>()
const {
129 "the info::device::version info descriptor can only "
130 "be queried with an OpenCL backend");
132 auto Devices = MKernelBundleImpl->get_devices();
133 if (Devices.empty()) {
134 return "No available device";
142 typename info::device::backend_version::return_type
143 kernel_impl::get_backend_info<info::device::backend_version>()
const {
146 "the info::device::backend_version info descriptor "
147 "can only be queried with a Level Zero backend");
bool isCreatedFromSource() const
Check if kernel was created from a program that had been created from source.
const PluginPtr & getPlugin() const
kernel_impl(ur_kernel_handle_t Kernel, ContextImplPtr Context, KernelBundleImplPtr KernelBundleImpl, const KernelArgMask *ArgMask=nullptr)
Constructs a SYCL kernel instance from a UrKernel.
The SYCL device class encapsulates a single SYCL device on which kernels may be executed.
detail::is_device_info_desc< Param >::return_type get_info() const
Queries this SYCL device for information requested by the template parameter param.
Objects of the class identify kernel is some kernel_bundle related APIs.
#define __SYCL_REPORT_EXCEPTION_TO_STREAM(str, e)
std::shared_ptr< device_image_impl > DeviceImageImplPtr
std::vector< bool > KernelArgMask
std::shared_ptr< sycl::detail::context_impl > ContextImplPtr
std::function< int(const sycl::device &)> DSelectorInvocableType
device select_device(const DSelectorInvocableType &DeviceSelectorInvocable)
std::shared_ptr< detail::kernel_bundle_impl > KernelBundleImplPtr
int default_selector_v(const device &dev)
std::error_code make_error_code(sycl::errc E) noexcept
Constructs an error code using e and sycl_category()
bool any_of(const simd_mask< _Tp, _Abi > &) noexcept