DPC++ Runtime
Runtime libraries for oneAPI DPC++
kernel_bundle.cpp
Go to the documentation of this file.
1 //==------- kernel_bundle.cpp - SYCL kernel_bundle and free functions ------==//
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 
14 
15 #include <set>
16 
17 namespace sycl {
18 inline namespace _V1 {
19 
20 kernel_id::kernel_id(const char *Name)
21  : impl(std::make_shared<detail::kernel_id_impl>(Name)) {}
22 
23 const char *kernel_id::get_name() const noexcept { return impl->get_name(); }
24 
25 namespace detail {
26 
30 
31 bool device_image_plain::has_kernel(const kernel_id &KernelID) const noexcept {
32  return impl->has_kernel(KernelID);
33 }
34 
35 bool device_image_plain::has_kernel(const kernel_id &KernelID,
36  const device &Dev) const noexcept {
37  return impl->has_kernel(KernelID, Dev);
38 }
39 
40 pi_native_handle device_image_plain::getNative() const {
41  return impl->getNative();
42 }
43 
47 
48 bool kernel_bundle_plain::empty() const noexcept { return impl->empty(); }
49 
50 backend kernel_bundle_plain::get_backend() const noexcept {
51  return impl->get_backend();
52 }
53 
54 context kernel_bundle_plain::get_context() const noexcept {
55  return impl->get_context();
56 }
57 
58 std::vector<device> kernel_bundle_plain::get_devices() const noexcept {
59  return impl->get_devices();
60 }
61 
62 std::vector<kernel_id> kernel_bundle_plain::get_kernel_ids() const {
63  return impl->get_kernel_ids();
64 }
65 
66 bool kernel_bundle_plain::contains_specialization_constants() const noexcept {
67  return impl->contains_specialization_constants();
68 }
69 
70 bool kernel_bundle_plain::native_specialization_constant() const noexcept {
71  return impl->native_specialization_constant();
72 }
73 
74 kernel kernel_bundle_plain::get_kernel(const kernel_id &KernelID) const {
75  return impl->get_kernel(KernelID, impl);
76 }
77 
78 const device_image_plain *kernel_bundle_plain::begin() const {
79  return impl->begin();
80 }
81 
82 const device_image_plain *kernel_bundle_plain::end() const {
83  return impl->end();
84 }
85 
86 bool kernel_bundle_plain::has_kernel(const kernel_id &KernelID) const noexcept {
87  return impl->has_kernel(KernelID);
88 }
89 
90 bool kernel_bundle_plain::has_kernel(const kernel_id &KernelID,
91  const device &Dev) const noexcept {
92  return impl->has_kernel(KernelID, Dev);
93 }
94 
95 bool kernel_bundle_plain::has_specialization_constant_impl(
96  const char *SpecName) const noexcept {
97  return impl->has_specialization_constant(SpecName);
98 }
99 
100 void kernel_bundle_plain::set_specialization_constant_impl(
101  const char *SpecName, void *Value, size_t Size) noexcept {
102  impl->set_specialization_constant_raw_value(SpecName, Value, Size);
103 }
104 
105 void kernel_bundle_plain::get_specialization_constant_impl(
106  const char *SpecName, void *Value) const noexcept {
107  impl->get_specialization_constant_raw_value(SpecName, Value);
108 }
109 
110 bool kernel_bundle_plain::is_specialization_constant_set(
111  const char *SpecName) const noexcept {
112  return impl->is_specialization_constant_set(SpecName);
113 }
114 
115 bool kernel_bundle_plain::ext_oneapi_has_kernel(const std::string &name) {
116  return impl->ext_oneapi_has_kernel(name);
117 }
118 
119 kernel kernel_bundle_plain::ext_oneapi_get_kernel(const std::string &name) {
120  return impl->ext_oneapi_get_kernel(name, impl);
121 }
122 
126 
127 const std::vector<device>
128 removeDuplicateDevices(const std::vector<device> &Devs) {
129  std::vector<device> UniqueDevices;
130 
131  // Building a new vector with unique elements and keep original order
132  std::unordered_set<device> UniqueDeviceSet;
133  for (const device &Dev : Devs)
134  if (UniqueDeviceSet.insert(Dev).second)
135  UniqueDevices.push_back(Dev);
136 
137  return UniqueDevices;
138 }
139 
140 kernel_id get_kernel_id_impl(std::string KernelName) {
141  return detail::ProgramManager::getInstance().getSYCLKernelID(KernelName);
142 }
143 
145 get_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
146  bundle_state State) {
147  return std::make_shared<detail::kernel_bundle_impl>(Ctx, Devs, State);
148 }
149 
151 get_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
152  const std::vector<kernel_id> &KernelIDs,
153  bundle_state State) {
154  return std::make_shared<detail::kernel_bundle_impl>(Ctx, Devs, KernelIDs,
155  State);
156 }
157 
159 get_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
160  bundle_state State, const DevImgSelectorImpl &Selector) {
161  return std::make_shared<detail::kernel_bundle_impl>(Ctx, Devs, Selector,
162  State);
163 }
164 
167  const std::vector<device> &Devs) {
168  return std::make_shared<detail::kernel_bundle_impl>(Ctx, Devs);
169 }
170 
171 std::shared_ptr<detail::kernel_bundle_impl>
172 join_impl(const std::vector<detail::KernelBundleImplPtr> &Bundles,
173  bundle_state State) {
174  return std::make_shared<detail::kernel_bundle_impl>(Bundles, State);
175 }
176 
177 bool has_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
178  bundle_state State) {
179  // Check that all requested devices are associated with the context
180  const bool AllDevicesInTheContext = checkAllDevicesAreInContext(Devs, Ctx);
181  if (Devs.empty() || !AllDevicesInTheContext)
182  throw sycl::exception(make_error_code(errc::invalid),
183  "Not all devices are associated with the context or "
184  "vector of devices is empty");
185 
186  if (bundle_state::input == State &&
187  !checkAllDevicesHaveAspect(Devs, aspect::online_compiler))
188  return false;
189  if (bundle_state::object == State &&
190  !checkAllDevicesHaveAspect(Devs, aspect::online_linker))
191  return false;
192 
193  const std::vector<device_image_plain> DeviceImages =
194  detail::ProgramManager::getInstance()
195  .getSYCLDeviceImagesWithCompatibleState(Ctx, Devs, State);
196 
197  return (bool)DeviceImages.size();
198 }
199 
200 bool has_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
201  const std::vector<kernel_id> &KernelIds,
202  bundle_state State) {
203  // Check that all requested devices are associated with the context
204  const bool AllDevicesInTheContext = checkAllDevicesAreInContext(Devs, Ctx);
205 
206  if (Devs.empty() || !AllDevicesInTheContext)
207  throw sycl::exception(make_error_code(errc::invalid),
208  "Not all devices are associated with the context or "
209  "vector of devices is empty");
210 
211  bool DeviceHasRequireAspectForState = true;
212  if (bundle_state::input == State) {
213  DeviceHasRequireAspectForState =
214  std::all_of(Devs.begin(), Devs.end(), [](const device &Dev) {
215  return Dev.has(aspect::online_compiler);
216  });
217  } else if (bundle_state::object == State) {
218  DeviceHasRequireAspectForState =
219  std::all_of(Devs.begin(), Devs.end(), [](const device &Dev) {
220  return Dev.has(aspect::online_linker);
221  });
222  }
223 
224  if (!DeviceHasRequireAspectForState)
225  return false;
226 
227  const std::vector<device_image_plain> DeviceImages =
228  detail::ProgramManager::getInstance()
229  .getSYCLDeviceImagesWithCompatibleState(Ctx, Devs, State);
230 
231  std::set<kernel_id, LessByNameComp> CombinedKernelIDs;
232  for (const device_image_plain &DeviceImage : DeviceImages) {
233  const std::shared_ptr<device_image_impl> &DeviceImageImpl =
234  getSyclObjImpl(DeviceImage);
235 
236  CombinedKernelIDs.insert(DeviceImageImpl->get_kernel_ids_ptr()->begin(),
237  DeviceImageImpl->get_kernel_ids_ptr()->end());
238  }
239 
240  const bool AllKernelIDsRepresented =
241  std::all_of(KernelIds.begin(), KernelIds.end(),
242  [&CombinedKernelIDs](const kernel_id &KernelID) {
243  return CombinedKernelIDs.count(KernelID);
244  });
245 
246  return AllKernelIDsRepresented;
247 }
248 
249 std::shared_ptr<detail::kernel_bundle_impl>
251  const std::vector<device> &Devs, const property_list &PropList) {
252  return std::make_shared<detail::kernel_bundle_impl>(
253  InputBundle, Devs, PropList, bundle_state::object);
254 }
255 
256 std::shared_ptr<detail::kernel_bundle_impl>
257 link_impl(const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles,
258  const std::vector<device> &Devs, const property_list &PropList) {
259  return std::make_shared<detail::kernel_bundle_impl>(ObjectBundles, Devs,
260  PropList);
261 }
262 
263 std::shared_ptr<detail::kernel_bundle_impl>
265  const std::vector<device> &Devs, const property_list &PropList) {
266  return std::make_shared<detail::kernel_bundle_impl>(
267  InputBundle, Devs, PropList, bundle_state::executable);
268 }
269 
270 // This function finds intersection of associated devices in common for all
271 // bundles
272 std::vector<sycl::device> find_device_intersection(
273  const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles) {
274  std::vector<sycl::device> IntersectDevices;
275  std::vector<unsigned int> DevsCounters;
276  std::map<device, unsigned int, LessByHash<device>> DevCounters;
277  for (const sycl::kernel_bundle<bundle_state::object> &ObjectBundle :
278  ObjectBundles)
279  // Increment counter in "DevCounters" each time a device is seen
280  for (const sycl::device &Device : ObjectBundle.get_devices())
281  DevCounters[Device]++;
282 
283  // If some device counter is less than ObjectBundles.size() then some bundle
284  // doesn't have it - do not add such a device to the final result
285  for (const std::pair<const device, unsigned int> &It : DevCounters)
286  if (ObjectBundles.size() == It.second)
287  IntersectDevices.push_back(It.first);
288 
289  return IntersectDevices;
290 }
291 
292 } // namespace detail
293 
297 
298 std::vector<kernel_id> get_kernel_ids() {
299  return detail::ProgramManager::getInstance().getAllSYCLKernelIDs();
300 }
301 
302 bool is_compatible(const std::vector<kernel_id> &KernelIDs, const device &Dev) {
303  if (KernelIDs.empty())
304  return true;
305  // TODO: also need to check that the architecture specified by the
306  // "-fsycl-targets" flag matches the device when we are able to get the
307  // device's arch.
308  auto doesImageTargetMatchDevice = [](const device &Dev,
309  const detail::RTDeviceBinaryImage &Img) {
310  const char *Target = Img.getRawData().DeviceTargetSpec;
311  auto BE = Dev.get_backend();
312  // ESIMD emulator backend is only compatible with esimd kernels.
313  if (BE == sycl::backend::ext_intel_esimd_emulator) {
314  pi_device_binary_property Prop = Img.getProperty("isEsimdImage");
315  return (Prop && (detail::DeviceBinaryProperty(Prop).asUint32() != 0));
316  }
317  if (strcmp(Target, __SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64) == 0) {
318  return (BE == sycl::backend::opencl ||
319  BE == sycl::backend::ext_oneapi_level_zero);
320  } else if (strcmp(Target, __SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64_X86_64) ==
321  0) {
322  return Dev.is_cpu();
323  } else if (strcmp(Target, __SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64_GEN) ==
324  0) {
325  return Dev.is_gpu() && (BE == sycl::backend::opencl ||
326  BE == sycl::backend::ext_oneapi_level_zero);
327  } else if (strcmp(Target, __SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64_FPGA) ==
328  0) {
329  return Dev.is_accelerator();
330  } else if (strcmp(Target, __SYCL_PI_DEVICE_BINARY_TARGET_NVPTX64) == 0) {
331  return BE == sycl::backend::ext_oneapi_cuda;
332  } else if (strcmp(Target, __SYCL_PI_DEVICE_BINARY_TARGET_AMDGCN) == 0) {
333  return BE == sycl::backend::ext_oneapi_hip;
334  }
335 
336  return false;
337  };
338 
339  // One kernel may be contained in several binary images depending on the
340  // number of targets. This kernel is compatible with the device if there is
341  // at least one image (containing this kernel) whose aspects are supported by
342  // the device and whose target matches the device.
343  for (const auto &KernelID : KernelIDs) {
344  std::set<detail::RTDeviceBinaryImage *> BinImages =
345  detail::ProgramManager::getInstance().getRawDeviceImages({KernelID});
346 
347  if (std::none_of(BinImages.begin(), BinImages.end(),
348  [&](const detail::RTDeviceBinaryImage *Img) {
349  return doesDevSupportDeviceRequirements(Dev, *Img) &&
350  doesImageTargetMatchDevice(Dev, *Img);
351  }))
352  return false;
353  }
354 
355  return true;
356 }
357 
359 // * kernel_compiler extension *
361 namespace ext::oneapi::experimental {
362 
366 
368 // syclex::is_source_kernel_bundle_supported
371  // at the moment, OpenCL is the only language supported
372  // and it's support is limited to the opencl and level_zero backends.
373  bool BE_Acceptable = (BE == sycl::backend::ext_oneapi_level_zero) ||
374  (BE == sycl::backend::opencl);
375  if ((Language == source_language::opencl) && BE_Acceptable) {
377  }
378 
379  // otherwise
380  return false;
381 }
382 
384 // syclex::create_kernel_bundle_from_source
387  source_language Language,
388  const std::string &Source) {
389  // TODO: if we later support a "reason" why support isn't present
390  // (like a missing shared library etc.) it'd be nice to include it in
391  // the exception message here.
392  backend BE = SyclContext.get_backend();
393  if (!is_source_kernel_bundle_supported(BE, Language))
394  throw sycl::exception(make_error_code(errc::invalid),
395  "kernel_bundle creation from source not supported");
396 
397  std::shared_ptr<kernel_bundle_impl> KBImpl =
398  std::make_shared<kernel_bundle_impl>(SyclContext, Language, Source);
399  return sycl::detail::createSyclObjFromImpl<source_kb>(KBImpl);
400 }
401 
403 // syclex::detail::build_from_source(source_kb) => exe_kb
405 namespace detail {
406 
408  const std::vector<device> &Devices,
409  const std::vector<std::string> &BuildOptions,
410  std::string *LogPtr) {
411  std::vector<device> UniqueDevices =
413  std::shared_ptr<kernel_bundle_impl> sourceImpl = getSyclObjImpl(SourceKB);
414  std::shared_ptr<kernel_bundle_impl> KBImpl =
415  sourceImpl->build_from_source(UniqueDevices, BuildOptions, LogPtr);
416  return sycl::detail::createSyclObjFromImpl<exe_kb>(KBImpl);
417 }
418 
419 } // namespace detail
420 } // namespace ext::oneapi::experimental
421 
422 } // namespace _V1
423 } // namespace sycl
sycl::_V1::property_list
Objects of the property_list class are containers for the SYCL properties.
Definition: property_list.hpp:31
sycl::_V1::ext::oneapi::experimental::detail::build_from_source
exe_kb build_from_source(source_kb &SourceKB, const std::vector< device > &Devices, const std::vector< std::string > &BuildOptions, std::string *LogPtr)
Definition: kernel_bundle.cpp:407
sycl::_V1::detail::checkAllDevicesAreInContext
static bool checkAllDevicesAreInContext(const std::vector< device > &Devices, const context &Context)
Definition: kernel_bundle_impl.hpp:33
sycl::_V1::backend
backend
Definition: backend_types.hpp:18
sycl::_V1::make_error_code
std::error_code make_error_code(sycl::errc E) noexcept
Constructs an error code using e and sycl_category()
Definition: exception.cpp:94
sycl::_V1::kernel_id::kernel_id
kernel_id()=delete
sycl::_V1::detail::get_kernel_id_impl
kernel_id get_kernel_id_impl(std::string KernelName)
Definition: kernel_bundle.cpp:140
sycl::_V1::ext::oneapi::experimental::is_source_kernel_bundle_supported
bool is_source_kernel_bundle_supported(backend BE, source_language Language)
Definition: kernel_bundle.cpp:370
sycl::_V1::detail::join_impl
std::shared_ptr< detail::kernel_bundle_impl > join_impl(const std::vector< detail::KernelBundleImplPtr > &Bundles, bundle_state State)
Definition: kernel_bundle.cpp:172
sycl
Definition: access.hpp:18
sycl::_V1::detail::find_device_intersection
std::vector< sycl::device > find_device_intersection(const std::vector< kernel_bundle< bundle_state::object >> &ObjectBundles)
Definition: kernel_bundle.cpp:272
kernel_id_impl.hpp
noexcept
_Abi const simd< _Tp, _Abi > & noexcept
Definition: simd.hpp:1324
sycl::_V1::detail::KernelBundleImplPtr
std::shared_ptr< detail::kernel_bundle_impl > KernelBundleImplPtr
Definition: kernel_bundle.hpp:150
__SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64_FPGA
#define __SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64_FPGA
Definition: pi.h:907
sycl::_V1::device::get_backend
backend get_backend() const noexcept
Returns the backend associated with this device.
Definition: device.cpp:205
sycl::_V1::detail::link_impl
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)
Definition: kernel_bundle.cpp:257
sycl::_V1::kernel
Provides an abstraction of a SYCL kernel.
Definition: kernel.hpp:74
sycl::_V1::ext::oneapi::experimental::source_language
source_language
Definition: kernel_bundle_enums.hpp:23
sycl::_V1::detail::DeviceBinaryProperty
Definition: device_binary_image.hpp:64
sycl::_V1::detail::get_empty_interop_kernel_bundle_impl
detail::KernelBundleImplPtr get_empty_interop_kernel_bundle_impl(const context &Ctx, const std::vector< device > &Devs)
Definition: kernel_bundle.cpp:166
sycl::_V1::detail::has_kernel_bundle_impl
bool has_kernel_bundle_impl(const context &Ctx, const std::vector< device > &Devs, bundle_state State)
Definition: kernel_bundle.cpp:177
sycl::_V1::detail::compile_impl
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)
Definition: kernel_bundle.cpp:250
sycl::_V1::device::is_gpu
bool is_gpu() const
Check if device is a GPU device.
Definition: device.cpp:87
__SYCL_PI_DEVICE_BINARY_TARGET_AMDGCN
#define __SYCL_PI_DEVICE_BINARY_TARGET_AMDGCN
Definition: pi.h:910
sycl::_V1::device::is_accelerator
bool is_accelerator() const
Check if device is an accelerator device.
Definition: device.cpp:89
sycl::_V1::kernel_id
Objects of the class identify kernel is some kernel_bundle related APIs.
Definition: kernel_bundle.hpp:56
sycl::_V1::exception
Definition: exception.hpp:68
sycl::_V1::kernel_bundle
The kernel_bundle class represents collection of device images in a particular state.
Definition: kernel.hpp:32
sycl::_V1::ext::oneapi::experimental::detail::OpenCLC_Compilation_Available
bool OpenCLC_Compilation_Available()
Definition: kernel_compiler_opencl.cpp:81
sycl::_V1::device
The SYCL device class encapsulates a single SYCL device on which kernels may be executed.
Definition: device.hpp:59
none_of
bool none_of(const simd_mask< _Tp, _Abi > &) noexcept
_pi_device_binary_property_struct
Definition: pi.h:855
sycl::_V1::context::get_backend
backend get_backend() const noexcept
Returns the backend associated with this context.
Definition: context.cpp:133
program_manager.hpp
pi_native_handle
uintptr_t pi_native_handle
Definition: pi.h:198
sycl::_V1::detail::removeDuplicateDevices
const std::vector< device > removeDuplicateDevices(const std::vector< device > &Devs)
Definition: kernel_bundle.cpp:128
sycl::_V1::device::get_devices
static std::vector< device > get_devices(info::device_type deviceType=info::device_type::all)
Query available SYCL devices.
Definition: device.cpp:52
all_of
bool all_of(const simd_mask< _Tp, _Abi > &) noexcept
sycl::_V1::ext::oneapi::experimental::create_kernel_bundle_from_source
kernel_bundle< bundle_state::ext_oneapi_source > create_kernel_bundle_from_source(const context &SyclContext, sycl::ext::oneapi::experimental::source_language Language, const std::string &Source)
Definition: kernel_bundle.cpp:386
sycl::_V1::detail::device_image_plain
Definition: kernel_bundle.hpp:88
sycl::_V1::ext::oneapi::experimental::detail::BuildOptions
@ BuildOptions
Definition: property.hpp:221
kernel_compiler_opencl.hpp
std
Definition: accessor.hpp:4171
device_binary_image.hpp
sycl::_V1::device::is_cpu
bool is_cpu() const
Check if device is a CPU device.
Definition: device.cpp:85
sycl::_V1::ext::oneapi::experimental::kernel_bundle_impl
sycl::detail::kernel_bundle_impl kernel_bundle_impl
Definition: kernel_bundle.cpp:365
kernel_bundle_impl.hpp
sycl::_V1::detail::get_kernel_bundle_impl
detail::KernelBundleImplPtr get_kernel_bundle_impl(const context &Ctx, const std::vector< device > &Devs, bundle_state State)
Definition: kernel_bundle.cpp:145
sycl::_V1::get_kernel_ids
std::vector< kernel_id > get_kernel_ids()
Definition: kernel_bundle.cpp:298
__SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64
#define __SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64
SPIR-V 64-bit image <-> "spir64", 64-bit OpenCL device.
Definition: pi.h:902
__SYCL_PI_DEVICE_BINARY_TARGET_NVPTX64
#define __SYCL_PI_DEVICE_BINARY_TARGET_NVPTX64
PTX 64-bit image <-> "nvptx64", 64-bit NVIDIA PTX device.
Definition: pi.h:909
__SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64_GEN
#define __SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64_GEN
Definition: pi.h:906
sycl::_V1::bundle_state
bundle_state
Definition: kernel_bundle_enums.hpp:14
sycl::_V1::is_compatible
bool is_compatible(const std::vector< kernel_id > &KernelIDs, const device &Dev)
Definition: kernel_bundle.cpp:302
sycl::_V1::detail::checkAllDevicesHaveAspect
static bool checkAllDevicesHaveAspect(const std::vector< device > &Devices, aspect Aspect)
Definition: kernel_bundle_impl.hpp:41
__SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64_X86_64
#define __SYCL_PI_DEVICE_BINARY_TARGET_SPIRV64_X86_64
Device-specific binary images produced from SPIR-V 64-bit <-> various "spir64_*" triples for specific...
Definition: pi.h:905
sycl::_V1::detail::RTDeviceBinaryImage
Definition: device_binary_image.hpp:82
sycl::_V1::detail::build_impl
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)
Definition: kernel_bundle.cpp:264
sycl::_V1::detail::getSyclObjImpl
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: impl_utils.hpp:30
sycl::_V1::context
The context class represents a SYCL context on which kernel functions may be executed.
Definition: context.hpp:51
sycl::_V1::detail::DevImgSelectorImpl
std::function< bool(const detail::DeviceImageImplPtr &DevImgImpl)> DevImgSelectorImpl
Definition: kernel_bundle.hpp:551