DPC++ Runtime
Runtime libraries for oneAPI DPC++
kernel_impl.hpp
Go to the documentation of this file.
1 //==------- kernel_impl.hpp --- SYCL kernel implementation -----------------==//
2 //
3 // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4 // See https://llvm.org/LICENSE.txt for license information.
5 // SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6 //
7 //===----------------------------------------------------------------------===//
8 
9 #pragma once
10 
11 #include <detail/context_impl.hpp>
12 #include <detail/device_impl.hpp>
13 #include <detail/kernel_info.hpp>
14 #include <sycl/detail/common.hpp>
15 #include <sycl/detail/pi.h>
16 #include <sycl/detail/pi.hpp>
17 #include <sycl/device.hpp>
18 #include <sycl/info/info_desc.hpp>
19 
20 #include <cassert>
21 #include <memory>
22 
23 namespace sycl {
25 namespace detail {
26 // Forward declaration
27 class program_impl;
28 class kernel_bundle_impl;
29 
30 using ContextImplPtr = std::shared_ptr<context_impl>;
31 using ProgramImplPtr = std::shared_ptr<program_impl>;
32 using KernelBundleImplPtr = std::shared_ptr<kernel_bundle_impl>;
33 class kernel_impl {
34 public:
44  kernel_impl(RT::PiKernel Kernel, ContextImplPtr Context,
45  KernelBundleImplPtr KernelBundleImpl);
46 
60  kernel_impl(RT::PiKernel Kernel, ContextImplPtr ContextImpl,
61  ProgramImplPtr ProgramImpl, bool IsCreatedFromSource,
62  KernelBundleImplPtr KernelBundleImpl);
63 
70  kernel_impl(RT::PiKernel Kernel, ContextImplPtr ContextImpl,
71  DeviceImageImplPtr DeviceImageImpl,
72  KernelBundleImplPtr KernelBundleImpl);
73 
78  kernel_impl(ContextImplPtr Context, ProgramImplPtr ProgramImpl);
79 
80  // This section means the object is non-movable and non-copyable
81  // There is no need of move and copy constructors in kernel_impl.
82  // If they need to be added, piKernelRetain method for MKernel
83  // should be present.
84  kernel_impl(const kernel_impl &) = delete;
85  kernel_impl(kernel_impl &&) = delete;
86  kernel_impl &operator=(const kernel_impl &) = delete;
87  kernel_impl &operator=(kernel_impl &&) = delete;
88 
89  ~kernel_impl();
90 
98  cl_kernel get() const {
99  if (is_host()) {
100  throw invalid_object_error(
101  "This instance of kernel doesn't support OpenCL interoperability.",
102  PI_ERROR_INVALID_KERNEL);
103  }
105  return pi::cast<cl_kernel>(MKernel);
106  }
107 
111  bool is_host() const { return MContext->is_host(); }
112 
113  const plugin &getPlugin() const { return MContext->getPlugin(); }
114 
119  template <typename Param> typename Param::return_type get_info() const;
120 
126  template <typename Param>
127  typename Param::return_type get_info(const device &Device) const;
128 
136  template <typename Param>
137  typename Param::return_type get_info(const device &Device,
138  const range<3> &WGSize) const;
139 
143  RT::PiKernel &getHandleRef() { return MKernel; }
148  const RT::PiKernel &getHandleRef() const { return MKernel; }
149 
154  bool isCreatedFromSource() const;
155 
156  const DeviceImageImplPtr &getDeviceImage() const { return MDeviceImageImpl; }
157 
159  const plugin &Plugin = MContext->getPlugin();
160 
161  if (Plugin.getBackend() == backend::opencl)
162  Plugin.call<PiApiKind::piKernelRetain>(MKernel);
163 
164  pi_native_handle NativeKernel = 0;
165  Plugin.call<PiApiKind::piextKernelGetNativeHandle>(MKernel, &NativeKernel);
166 
167  return NativeKernel;
168  }
169 
170  KernelBundleImplPtr get_kernel_bundle() const { return MKernelBundleImpl; }
171 
172  bool isInterop() const { return MIsInterop; }
173 
174  ProgramImplPtr getProgramImpl() const { return MProgramImpl; }
175 
177  return MNoncacheableEnqueueMutex;
178  }
179 
180 private:
181  RT::PiKernel MKernel;
182  const ContextImplPtr MContext;
183  const ProgramImplPtr MProgramImpl;
184  bool MCreatedFromSource = true;
185  const DeviceImageImplPtr MDeviceImageImpl;
186  const KernelBundleImplPtr MKernelBundleImpl;
187  bool MIsInterop = false;
188  std::mutex MNoncacheableEnqueueMutex;
189 };
190 
191 template <typename Param>
192 inline typename Param::return_type kernel_impl::get_info() const {
193  static_assert(is_kernel_info_desc<Param>::value,
194  "Invalid kernel information descriptor");
195  if (is_host()) {
196  // TODO implement
197  assert(0 && "Not implemented");
198  }
199  return get_kernel_info<Param>(this->getHandleRef(), getPlugin());
200 }
201 
202 template <>
203 inline context kernel_impl::get_info<info::kernel::context>() const {
204  return createSyclObjFromImpl<context>(MContext);
205 }
206 
207 template <typename Param>
208 inline typename Param::return_type
209 kernel_impl::get_info(const device &Device) const {
210  if (is_host()) {
211  return get_kernel_device_specific_info_host<Param>(Device);
212  }
213  return get_kernel_device_specific_info<Param>(
214  this->getHandleRef(), getSyclObjImpl(Device)->getHandleRef(),
215  getPlugin());
216 }
217 
218 template <typename Param>
219 inline typename Param::return_type
220 kernel_impl::get_info(const device &Device,
221  const sycl::range<3> &WGSize) const {
222  if (is_host()) {
223  throw runtime_error("Sub-group feature is not supported on HOST device.",
224  PI_ERROR_INVALID_DEVICE);
225  }
226  return get_kernel_device_specific_info_with_input<Param>(
227  this->getHandleRef(), getSyclObjImpl(Device)->getHandleRef(), WGSize,
228  getPlugin());
229 }
230 
231 } // namespace detail
232 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
233 } // namespace sycl
pi.h
context_impl.hpp
sycl::_V1::detail::ContextImplPtr
std::shared_ptr< sycl::detail::context_impl > ContextImplPtr
Definition: event_impl.hpp:30
piextKernelGetNativeHandle
pi_result piextKernelGetNativeHandle(pi_kernel kernel, pi_native_handle *nativeHandle)
Gets the native handle of a PI kernel object.
Definition: pi_esimd_emulator.cpp:1879
device.hpp
sycl::_V1::detail::DeviceImageImplPtr
std::shared_ptr< device_image_impl > DeviceImageImplPtr
Definition: kernel_bundle.hpp:72
__SYCL_INLINE_VER_NAMESPACE
#define __SYCL_INLINE_VER_NAMESPACE(X)
Definition: defines_elementary.hpp:11
sycl::_V1::detail::kernel_impl::getProgramImpl
ProgramImplPtr getProgramImpl() const
Definition: kernel_impl.hpp:174
sycl::_V1::detail::ProgramImplPtr
std::shared_ptr< program_impl > ProgramImplPtr
Definition: kernel_impl.hpp:31
sycl::_V1::detail::kernel_impl
Definition: kernel_impl.hpp:33
sycl
---— Error handling, matching OpenCL plugin semantics.
Definition: access.hpp:14
device_impl.hpp
sycl::_V1::detail::plugin::getBackend
backend getBackend(void) const
Definition: plugin.hpp:229
sycl::_V1::detail::KernelBundleImplPtr
std::shared_ptr< detail::kernel_bundle_impl > KernelBundleImplPtr
Definition: kernel_bundle.hpp:138
pi.hpp
sycl::_V1::detail::kernel_impl::getHandleRef
RT::PiKernel & getHandleRef()
Get a reference to a raw kernel object.
Definition: kernel_impl.hpp:143
sycl::_V1::detail::kernel_impl::getNative
pi_native_handle getNative() const
Definition: kernel_impl.hpp:158
sycl::_V1::detail::pi::getPlugin
const plugin & getPlugin()
Definition: pi.cpp:506
sycl::_V1::range< 3 >
sycl::_V1::detail::plugin
The plugin class provides a unified interface to the underlying low-level runtimes for the device-agn...
Definition: plugin.hpp:90
sycl::_V1::detail::kernel_impl::get_kernel_bundle
KernelBundleImplPtr get_kernel_bundle() const
Definition: kernel_impl.hpp:170
sycl::_V1::ext::oneapi::experimental::operator=
annotated_arg & operator=(annotated_arg &)=default
piKernelRetain
pi_result piKernelRetain(pi_kernel kernel)
Definition: pi_esimd_emulator.cpp:1398
sycl::_V1::detail::kernel_impl::get
cl_kernel get() const
Gets a valid OpenCL kernel handle.
Definition: kernel_impl.hpp:98
common.hpp
sycl::_V1::device
The SYCL device class encapsulates a single SYCL device on which kernels may be executed.
Definition: device.hpp:49
pi_native_handle
uintptr_t pi_native_handle
Definition: pi.h:133
sycl::_V1::detail::pi::PiKernel
::pi_kernel PiKernel
Definition: pi.hpp:131
sycl::_V1::detail::kernel_impl::getDeviceImage
const DeviceImageImplPtr & getDeviceImage() const
Definition: kernel_impl.hpp:156
kernel_info.hpp
sycl::_V1::detail::kernel_impl::isInterop
bool isInterop() const
Definition: kernel_impl.hpp:172
sycl::_V1::detail::plugin::call
void call(ArgsT... Args) const
Calls the API, traces the call, checks the result.
Definition: plugin.hpp:217
sycl::_V1::detail::kernel_impl::getHandleRef
const RT::PiKernel & getHandleRef() const
Get a constant reference to a raw kernel object.
Definition: kernel_impl.hpp:148
sycl::_V1::detail::is_kernel_info_desc
Definition: info_desc_helpers.hpp:23
info_desc.hpp
sycl::_V1::detail::kernel_impl::is_host
bool is_host() const
Check if the associated SYCL context is a SYCL host context.
Definition: kernel_impl.hpp:111
sycl::_V1::detail::kernel_impl::getNoncacheableEnqueueMutex
std::mutex & getNoncacheableEnqueueMutex()
Definition: kernel_impl.hpp:176
sycl::_V1::detail::getSyclObjImpl
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:300
sycl::_V1::context
The context class represents a SYCL context on which kernel functions may be executed.
Definition: context.hpp:41
sycl::_V1::detail::kernel_impl::getPlugin
const plugin & getPlugin() const
Definition: kernel_impl.hpp:113