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;
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 
176 private:
177  RT::PiKernel MKernel;
178  const ContextImplPtr MContext;
179  const ProgramImplPtr MProgramImpl;
180  bool MCreatedFromSource = true;
181  const DeviceImageImplPtr MDeviceImageImpl;
182  const KernelBundleImplPtr MKernelBundleImpl;
183  bool MIsInterop = false;
184 };
185 
186 template <typename Param>
187 inline typename Param::return_type kernel_impl::get_info() const {
188  static_assert(is_kernel_info_desc<Param>::value,
189  "Invalid kernel information descriptor");
190  if (is_host()) {
191  // TODO implement
192  assert(0 && "Not implemented");
193  }
194  return get_kernel_info<Param>(this->getHandleRef(), getPlugin());
195 }
196 
197 template <>
198 inline context kernel_impl::get_info<info::kernel::context>() const {
199  return createSyclObjFromImpl<context>(MContext);
200 }
201 
202 template <typename Param>
203 inline typename Param::return_type
204 kernel_impl::get_info(const device &Device) const {
205  if (is_host()) {
206  return get_kernel_device_specific_info_host<Param>(Device);
207  }
208  return get_kernel_device_specific_info<Param>(
209  this->getHandleRef(), getSyclObjImpl(Device)->getHandleRef(),
210  getPlugin());
211 }
212 
213 template <typename Param>
214 inline typename Param::return_type
215 kernel_impl::get_info(const device &Device,
216  const sycl::range<3> &WGSize) const {
217  if (is_host()) {
218  throw runtime_error("Sub-group feature is not supported on HOST device.",
219  PI_ERROR_INVALID_DEVICE);
220  }
221  return get_kernel_device_specific_info_with_input<Param>(
222  this->getHandleRef(), getSyclObjImpl(Device)->getHandleRef(), WGSize,
223  getPlugin());
224 }
225 
226 } // namespace detail
227 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
228 } // namespace sycl
The context class represents a SYCL context on which kernel functions may be executed.
Definition: context.hpp:41
const RT::PiKernel & getHandleRef() const
Get a constant reference to a raw kernel object.
kernel_impl(const kernel_impl &)=delete
kernel_impl & operator=(const kernel_impl &)=delete
kernel_impl & operator=(kernel_impl &&)=delete
Param::return_type get_info(const device &Device, const range< 3 > &WGSize) const
Query device-specific information from a kernel using the info::kernel_device_specific descriptor for...
bool is_host() const
Check if the associated SYCL context is a SYCL host context.
const plugin & getPlugin() const
kernel_impl(kernel_impl &&)=delete
ProgramImplPtr getProgramImpl() const
pi_native_handle getNative() const
RT::PiKernel & getHandleRef()
Get a reference to a raw kernel object.
const DeviceImageImplPtr & getDeviceImage() const
KernelBundleImplPtr get_kernel_bundle() const
cl_kernel get() const
Gets a valid OpenCL kernel handle.
Definition: kernel_impl.hpp:98
The plugin class provides a unified interface to the underlying low-level runtimes for the device-agn...
Definition: plugin.hpp:90
void call(ArgsT... Args) const
Calls the API, traces the call, checks the result.
Definition: plugin.hpp:217
backend getBackend(void) const
Definition: plugin.hpp:229
The SYCL device class encapsulates a single SYCL device on which kernels may be executed.
Definition: device.hpp:49
#define __SYCL_INLINE_VER_NAMESPACE(X)
::pi_kernel PiKernel
Definition: pi.hpp:116
const plugin & getPlugin()
Definition: pi.cpp:509
std::shared_ptr< device_image_impl > DeviceImageImplPtr
std::shared_ptr< program_impl > ProgramImplPtr
Definition: kernel_impl.hpp:31
std::shared_ptr< sycl::detail::context_impl > ContextImplPtr
Definition: event_impl.hpp:30
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:248
std::shared_ptr< detail::kernel_bundle_impl > KernelBundleImplPtr
---— Error handling, matching OpenCL plugin semantics.
Definition: access.hpp:14
uintptr_t pi_native_handle
Definition: pi.h:111
pi_result piKernelRetain(pi_kernel kernel)
pi_result piextKernelGetNativeHandle(pi_kernel kernel, pi_native_handle *nativeHandle)
Gets the native handle of a PI kernel object.
C++ wrapper of extern "C" PI interfaces.
@ Device