DPC++ Runtime
Runtime libraries for oneAPI Data Parallel C++
kernel_bundle.hpp
Go to the documentation of this file.
1 //==------- kernel_bundle.hpp - 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 
9 #pragma once
10 
11 #include <CL/sycl/context.hpp>
14 #include <CL/sycl/detail/pi.h>
15 #include <CL/sycl/detail/pi.hpp>
16 #include <CL/sycl/device.hpp>
17 #include <CL/sycl/kernel.hpp>
19 
20 #include <cassert>
21 #include <memory>
22 #include <set>
23 #include <vector>
24 
26 namespace sycl {
27 // Forward declaration
28 template <backend Backend> class backend_traits;
29 template <backend Backend, class SyclT>
30 auto get_native(const SyclT &Obj) -> backend_return_t<Backend, SyclT>;
31 
32 namespace detail {
33 class kernel_id_impl;
34 }
35 
39 class __SYCL_EXPORT kernel_id {
40 public:
41  kernel_id() = delete;
42 
44  const char *get_name() const noexcept;
45 
46  bool operator==(const kernel_id &RHS) const { return impl == RHS.impl; }
47 
48  bool operator!=(const kernel_id &RHS) const { return !(*this == RHS); }
49 
50 private:
51  kernel_id(const char *Name);
52 
53  kernel_id(const std::shared_ptr<detail::kernel_id_impl> &Impl)
54  : impl(std::move(Impl)) {}
55 
56  std::shared_ptr<detail::kernel_id_impl> impl;
57 
58  template <class Obj>
59  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
60 
61  template <class T>
62  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
63 };
64 
65 namespace detail {
66 class device_image_impl;
67 using DeviceImageImplPtr = std::shared_ptr<device_image_impl>;
68 
69 // The class is used as a base for device_image for "untemplating" public
70 // methods.
71 class __SYCL_EXPORT device_image_plain {
72 public:
74  : impl(std::move(Impl)) {}
75 
76  bool operator==(const device_image_plain &RHS) const {
77  return impl == RHS.impl;
78  }
79 
80  bool operator!=(const device_image_plain &RHS) const {
81  return !(*this == RHS);
82  }
83 
84  bool has_kernel(const kernel_id &KernelID) const noexcept;
85 
86  bool has_kernel(const kernel_id &KernelID, const device &Dev) const noexcept;
87 
88  pi_native_handle getNative() const;
89 
90 protected:
92 
93  template <class Obj>
94  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
95 
96  template <class T>
97  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
98 };
99 } // namespace detail
100 
102 template <sycl::bundle_state State>
104 public:
105  device_image() = delete;
106 
109  bool has_kernel(const kernel_id &KernelID) const noexcept {
110  return device_image_plain::has_kernel(KernelID);
111  }
112 
115  bool has_kernel(const kernel_id &KernelID, const device &Dev) const noexcept {
116  return device_image_plain::has_kernel(KernelID, Dev);
117  }
118 
119 private:
121  : device_image_plain(std::move(Impl)) {}
122 
123  template <class Obj>
124  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
125 
126  template <class T>
127  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
128 };
129 
130 namespace detail {
131 class kernel_bundle_impl;
132 using KernelBundleImplPtr = std::shared_ptr<detail::kernel_bundle_impl>;
133 
134 // The class is used as a base for kernel_bundle to "untemplate" it's methods
135 class __SYCL_EXPORT kernel_bundle_plain {
136 public:
138  : impl(std::move(Impl)) {}
139 
140  bool operator==(const kernel_bundle_plain &RHS) const {
141  return impl == RHS.impl;
142  }
143 
144  bool operator!=(const kernel_bundle_plain &RHS) const {
145  return !(*this == RHS);
146  }
147 
148  bool empty() const noexcept;
149 
150  backend get_backend() const noexcept;
151 
152  context get_context() const noexcept;
153 
154  std::vector<device> get_devices() const noexcept;
155 
156  bool has_kernel(const kernel_id &KernelID) const noexcept;
157 
158  bool has_kernel(const kernel_id &KernelID, const device &Dev) const noexcept;
159 
160  std::vector<kernel_id> get_kernel_ids() const;
161 
162  bool contains_specialization_constants() const noexcept;
163 
164  bool native_specialization_constant() const noexcept;
165 
166 protected:
167  // \returns a kernel object which represents the kernel identified by
168  // kernel_id passed
169  kernel get_kernel(const kernel_id &KernelID) const;
170 
171  // \returns an iterator to the first device image kernel_bundle contains
172  const device_image_plain *begin() const;
173 
174  // \returns an iterator to the last device image kernel_bundle contains
175  const device_image_plain *end() const;
176 
177  bool has_specialization_constant_impl(const char *SpecName) const noexcept;
178 
179  void set_specialization_constant_impl(const char *SpecName, void *Value,
180  size_t Size) noexcept;
181 
182  void get_specialization_constant_impl(const char *SpecName,
183  void *Value) const noexcept;
184 
185  bool is_specialization_constant_set(const char *SpecName) const noexcept;
186 
188 };
189 
190 } // namespace detail
191 
196 template <bundle_state State>
197 class kernel_bundle : public detail::kernel_bundle_plain {
198 public:
200 
201  kernel_bundle() = delete;
202 
204  bool empty() const noexcept { return kernel_bundle_plain::empty(); }
205 
207  backend get_backend() const noexcept {
208  return kernel_bundle_plain::get_backend();
209  }
210 
212  context get_context() const noexcept {
213  return kernel_bundle_plain::get_context();
214  }
215 
217  std::vector<device> get_devices() const noexcept {
218  return kernel_bundle_plain::get_devices();
219  }
220 
223  bool has_kernel(const kernel_id &KernelID) const noexcept {
224  return kernel_bundle_plain::has_kernel(KernelID);
225  }
226 
230  bool has_kernel(const kernel_id &KernelID, const device &Dev) const noexcept {
231  return kernel_bundle_plain::has_kernel(KernelID, Dev);
232  }
233 
235  std::vector<kernel_id> get_kernel_ids() const {
237  }
238 
241  bool contains_specialization_constants() const noexcept {
242  return kernel_bundle_plain::contains_specialization_constants();
243  }
244 
247  bool native_specialization_constant() const noexcept {
248  return kernel_bundle_plain::native_specialization_constant();
249  }
250 
253  template <bundle_state _State = State,
255  kernel get_kernel(const kernel_id &KernelID) const {
256  return detail::kernel_bundle_plain::get_kernel(KernelID);
257  }
258 
259  // This guard is needed because the libsycl.so can compiled with C++ <=14
260  // while the code requires C++17. This code is not supposed to be used by the
261  // libsycl.so so it should not be a problem.
262 #if __cplusplus > 201402L
263  template <auto &SpecName> bool has_specialization_constant() const noexcept {
266  const char *SpecSymName = detail::get_spec_constant_symbolic_ID<SpecName>();
267  return has_specialization_constant_impl(SpecSymName);
268  }
269 
273  template <auto &SpecName, bundle_state _State = State,
274  typename = detail::enable_if_t<_State == bundle_state::input>>
275  void set_specialization_constant(
276  typename std::remove_reference_t<decltype(SpecName)>::value_type Value) {
277  const char *SpecSymName = detail::get_spec_constant_symbolic_ID<SpecName>();
278  set_specialization_constant_impl(SpecSymName, &Value,
279  sizeof(decltype(Value)));
280  }
281 
284  template <auto &SpecName>
285  typename std::remove_reference_t<decltype(SpecName)>::value_type
286  get_specialization_constant() const {
287  const char *SpecSymName = detail::get_spec_constant_symbolic_ID<SpecName>();
288  if (!is_specialization_constant_set(SpecSymName))
289  return SpecName.getDefaultValue();
290 
291  using SCType =
292  typename std::remove_reference_t<decltype(SpecName)>::value_type;
293 
294  std::array<char *, sizeof(SCType)> RetValue;
295 
296  get_specialization_constant_impl(SpecSymName, RetValue.data());
297 
298  return *reinterpret_cast<SCType *>(RetValue.data());
299  }
300 #endif
301 
304  return reinterpret_cast<device_image_iterator>(
305  kernel_bundle_plain::begin());
306  }
307 
310  return reinterpret_cast<device_image_iterator>(kernel_bundle_plain::end());
311  }
312 
313  template <backend Backend>
314  __SYCL_DEPRECATED("Use SYCL 2020 sycl::get_native free function")
315  backend_return_t<Backend, kernel_bundle<State>> get_native() const {
316  return getNative<Backend>();
317  }
318 
319 private:
321  : kernel_bundle_plain(std::move(Impl)) {}
322 
323  template <class Obj>
324  friend decltype(Obj::impl) detail::getSyclObjImpl(const Obj &SyclObject);
325 
326  template <class T>
327  friend T detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
328 
329  template <backend Backend, class SyclT>
330  friend auto get_native(const SyclT &Obj) -> backend_return_t<Backend, SyclT>;
331 
332  template <backend Backend>
333  backend_return_t<Backend, kernel_bundle<State>> getNative() const {
334  // NOTE: implementation assumes that the return type is a
335  // derivative of std::vector.
336  backend_return_t<Backend, kernel_bundle<State>> ReturnValue;
337  ReturnValue.reserve(std::distance(begin(), end()));
338 
339  for (const device_image<State> &DevImg : *this) {
340  ReturnValue.push_back(
341  detail::pi::cast<typename decltype(ReturnValue)::value_type>(
342  DevImg.getNative()));
343  }
344 
345  return ReturnValue;
346  }
347 };
348 
350 // get_kernel_id API
352 
353 namespace detail {
354 // Internal non-template versions of get_kernel_id API which is used by public
355 // onces
356 __SYCL_EXPORT kernel_id get_kernel_id_impl(std::string KernelName);
357 } // namespace detail
358 
360 template <typename KernelName> kernel_id get_kernel_id() {
362  return detail::get_kernel_id_impl(KI::getName());
363 }
364 
366 __SYCL_EXPORT std::vector<kernel_id> get_kernel_ids();
367 
369 // get_kernel_bundle API
371 
372 namespace detail {
373 
374 // Internal non-template versions of get_kernel_bundle API which is used by
375 // public onces
376 __SYCL_EXPORT detail::KernelBundleImplPtr
377 get_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
378  bundle_state State);
379 
381  return [](device a, device b) { return a.getNative() != b.getNative(); };
382 }
383 
384 inline const std::vector<device>
385 removeDuplicateDevices(const std::vector<device> &Devs) {
386  auto compareDevices = getDeviceComparisonLambda();
387  std::set<device, decltype(compareDevices)> UniqueDeviceSet(
388  Devs.begin(), Devs.end(), compareDevices);
389  std::vector<device> UniqueDevices(UniqueDeviceSet.begin(),
390  UniqueDeviceSet.end());
391  return UniqueDevices;
392 }
393 
394 } // namespace detail
395 
400 template <bundle_state State>
402  const std::vector<device> &Devs) {
403  std::vector<device> UniqueDevices = detail::removeDuplicateDevices(Devs);
404 
406  detail::get_kernel_bundle_impl(Ctx, UniqueDevices, State);
407 
408  return detail::createSyclObjFromImpl<kernel_bundle<State>>(Impl);
409 }
410 
411 template <bundle_state State>
413  return get_kernel_bundle<State>(Ctx, Ctx.get_devices());
414 }
415 
416 namespace detail {
417 
418 // Internal non-template versions of get_kernel_bundle API which is used by
419 // public onces
420 __SYCL_EXPORT detail::KernelBundleImplPtr
421 get_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
422  const std::vector<kernel_id> &KernelIDs,
423  bundle_state State);
424 } // namespace detail
425 
434 template <bundle_state State>
435 kernel_bundle<State>
436 get_kernel_bundle(const context &Ctx, const std::vector<device> &Devs,
437  const std::vector<kernel_id> &KernelIDs) {
438  std::vector<device> UniqueDevices = detail::removeDuplicateDevices(Devs);
439 
441  detail::get_kernel_bundle_impl(Ctx, UniqueDevices, KernelIDs, State);
442  return detail::createSyclObjFromImpl<kernel_bundle<State>>(Impl);
443 }
444 
445 template <bundle_state State>
446 kernel_bundle<State>
447 get_kernel_bundle(const context &Ctx, const std::vector<kernel_id> &KernelIDs) {
448  return get_kernel_bundle<State>(Ctx, Ctx.get_devices(), KernelIDs);
449 }
450 
451 template <typename KernelName, bundle_state State>
453  return get_kernel_bundle<State>(Ctx, Ctx.get_devices(),
454  {get_kernel_id<KernelName>()});
455 }
456 
457 template <typename KernelName, bundle_state State>
459  const std::vector<device> &Devs) {
460  return get_kernel_bundle<State>(Ctx, Devs, {get_kernel_id<KernelName>()});
461 }
462 
463 namespace detail {
464 
465 // Stable selector function type for passing thru library boundaries
466 using DevImgSelectorImpl =
467  std::function<bool(const detail::DeviceImageImplPtr &DevImgImpl)>;
468 
469 // Internal non-template versions of get_kernel_bundle API which is used by
470 // public onces
471 __SYCL_EXPORT detail::KernelBundleImplPtr
472 get_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
473  bundle_state State, const DevImgSelectorImpl &Selector);
474 
475 // Internal non-template versions of get_empty_interop_kernel_bundle API which
476 // is used by public onces
477 __SYCL_EXPORT detail::KernelBundleImplPtr
479  const std::vector<device> &Devs);
480 
483 template <bundle_state State>
487  return detail::createSyclObjFromImpl<sycl::kernel_bundle<State>>(Impl);
488 }
489 } // namespace detail
490 
493 template <bundle_state State, typename SelectorT>
495  const std::vector<device> &Devs,
496  SelectorT Selector) {
497  std::vector<device> UniqueDevices = detail::removeDuplicateDevices(Devs);
498 
499  detail::DevImgSelectorImpl SelectorWrapper =
500  [Selector](const detail::DeviceImageImplPtr &DevImg) {
501  return Selector(
503  };
504 
506  Ctx, UniqueDevices, State, SelectorWrapper);
507 
508  return detail::createSyclObjFromImpl<sycl::kernel_bundle<State>>(Impl);
509 }
510 
511 template <bundle_state State, typename SelectorT>
512 kernel_bundle<State> get_kernel_bundle(const context &Ctx, SelectorT Selector) {
513  return get_kernel_bundle<State>(Ctx, Ctx.get_devices(), Selector);
514 }
515 
517 // has_kernel_bundle API
519 
520 namespace detail {
521 
522 __SYCL_EXPORT bool has_kernel_bundle_impl(const context &Ctx,
523  const std::vector<device> &Devs,
524  bundle_state State);
525 
526 __SYCL_EXPORT bool
527 has_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
528  const std::vector<kernel_id> &kernelIds,
529  bundle_state State);
530 } // namespace detail
531 
542 template <bundle_state State>
543 bool has_kernel_bundle(const context &Ctx, const std::vector<device> &Devs) {
544  return detail::has_kernel_bundle_impl(Ctx, Devs, State);
545 }
546 
547 template <bundle_state State>
548 bool has_kernel_bundle(const context &Ctx, const std::vector<device> &Devs,
549  const std::vector<kernel_id> &KernelIDs) {
550  return detail::has_kernel_bundle_impl(Ctx, Devs, KernelIDs, State);
551 }
552 
553 template <bundle_state State> bool has_kernel_bundle(const context &Ctx) {
554  return has_kernel_bundle<State>(Ctx, Ctx.get_devices());
555 }
556 
557 template <bundle_state State>
558 bool has_kernel_bundle(const context &Ctx,
559  const std::vector<kernel_id> &KernelIDs) {
560  return has_kernel_bundle<State>(Ctx, Ctx.get_devices(), KernelIDs);
561 }
562 
563 template <typename KernelName, bundle_state State>
564 bool has_kernel_bundle(const context &Ctx) {
565  return has_kernel_bundle<State>(Ctx, {get_kernel_id<KernelName>()});
566 }
567 
568 template <typename KernelName, bundle_state State>
569 bool has_kernel_bundle(const context &Ctx, const std::vector<device> &Devs) {
570  return has_kernel_bundle<State>(Ctx, Devs, {get_kernel_id<KernelName>()});
571 }
572 
574 // is_compatible API
576 
579 bool is_compatible(const std::vector<kernel_id> &KernelIDs, const device &Dev);
580 
581 template <typename KernelName> bool is_compatible(const device &Dev) {
582  return is_compatible({get_kernel_id<KernelName>()}, Dev);
583 }
584 
586 // join API
588 
589 namespace detail {
590 
591 // TODO: This is no longer in use. Remove when ABI break is allowed.
592 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
593 join_impl(const std::vector<detail::KernelBundleImplPtr> &Bundles);
594 
595 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
596 join_impl(const std::vector<detail::KernelBundleImplPtr> &Bundles,
597  bundle_state State);
598 }
599 
602 template <sycl::bundle_state State>
603 sycl::kernel_bundle<State>
604 join(const std::vector<sycl::kernel_bundle<State>> &Bundles) {
605  // Convert kernel_bundle<State> to impls to abstract template parameter away
606  std::vector<detail::KernelBundleImplPtr> KernelBundleImpls;
607  KernelBundleImpls.reserve(Bundles.size());
608  for (const sycl::kernel_bundle<State> &Bundle : Bundles)
609  KernelBundleImpls.push_back(detail::getSyclObjImpl(Bundle));
610 
611  std::shared_ptr<detail::kernel_bundle_impl> Impl =
612  detail::join_impl(KernelBundleImpls, State);
613  return detail::createSyclObjFromImpl<kernel_bundle<State>>(Impl);
614 }
615 
617 // compile API
619 
620 namespace detail {
621 
622 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
623 compile_impl(const kernel_bundle<bundle_state::input> &InputBundle,
624  const std::vector<device> &Devs, const property_list &PropList);
625 }
626 
631 inline kernel_bundle<bundle_state::object>
633  const std::vector<device> &Devs, const property_list &PropList = {}) {
634  std::vector<device> UniqueDevices = detail::removeDuplicateDevices(Devs);
635 
637  detail::compile_impl(InputBundle, UniqueDevices, PropList);
639  kernel_bundle<sycl::bundle_state::object>>(Impl);
640 }
641 
642 inline kernel_bundle<bundle_state::object>
644  const property_list &PropList = {}) {
645  return compile(InputBundle, InputBundle.get_devices(), PropList);
646 }
647 
649 // link API
651 
652 namespace detail {
653 __SYCL_EXPORT std::vector<sycl::device> find_device_intersection(
654  const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles);
655 
656 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
657 link_impl(const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles,
658  const std::vector<device> &Devs, const property_list &PropList);
659 } // namespace detail
660 
666 inline kernel_bundle<bundle_state::executable>
667 link(const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles,
668  const std::vector<device> &Devs, const property_list &PropList = {}) {
669  std::vector<device> UniqueDevices = detail::removeDuplicateDevices(Devs);
670 
672  detail::link_impl(ObjectBundles, UniqueDevices, PropList);
674  kernel_bundle<sycl::bundle_state::executable>>(Impl);
675 }
676 
677 inline kernel_bundle<bundle_state::executable>
679  const property_list &PropList = {}) {
680  return link(std::vector<kernel_bundle<bundle_state::object>>{ObjectBundle},
681  ObjectBundle.get_devices(), PropList);
682 }
683 
684 inline kernel_bundle<bundle_state::executable>
685 link(const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles,
686  const property_list &PropList = {}) {
687  std::vector<sycl::device> IntersectDevices =
688  find_device_intersection(ObjectBundles);
689  return link(ObjectBundles, IntersectDevices, PropList);
690 }
691 
692 inline kernel_bundle<bundle_state::executable>
694  const std::vector<device> &Devs, const property_list &PropList = {}) {
695  return link(std::vector<kernel_bundle<bundle_state::object>>{ObjectBundle},
696  Devs, PropList);
697 }
698 
700 // build API
702 
703 namespace detail {
704 __SYCL_EXPORT std::shared_ptr<detail::kernel_bundle_impl>
705 build_impl(const kernel_bundle<bundle_state::input> &InputBundle,
706  const std::vector<device> &Devs, const property_list &PropList);
707 }
708 
713 inline kernel_bundle<bundle_state::executable>
715  const std::vector<device> &Devs, const property_list &PropList = {}) {
716  std::vector<device> UniqueDevices = detail::removeDuplicateDevices(Devs);
717 
719  detail::build_impl(InputBundle, UniqueDevices, PropList);
721  kernel_bundle<sycl::bundle_state::executable>>(Impl);
722 }
723 
724 inline kernel_bundle<bundle_state::executable>
726  const property_list &PropList = {}) {
727  return build(InputBundle, InputBundle.get_devices(), PropList);
728 }
729 
730 } // namespace sycl
731 } // __SYCL_INLINE_NAMESPACE(cl)
732 
733 namespace std {
734 template <> struct hash<cl::sycl::kernel_id> {
735  size_t operator()(const cl::sycl::kernel_id &KernelID) const {
736  return hash<std::shared_ptr<cl::sycl::detail::kernel_id_impl>>()(
738  }
739 };
740 
741 template <cl::sycl::bundle_state State>
742 struct hash<cl::sycl::device_image<State>> {
743  size_t operator()(const cl::sycl::device_image<State> &DeviceImage) const {
744  return hash<std::shared_ptr<cl::sycl::detail::device_image_impl>>()(
745  cl::sycl::detail::getSyclObjImpl(DeviceImage));
746  }
747 };
748 
749 template <cl::sycl::bundle_state State>
750 struct hash<cl::sycl::kernel_bundle<State>> {
751  size_t operator()(const cl::sycl::kernel_bundle<State> &KernelBundle) const {
752  return hash<std::shared_ptr<cl::sycl::detail::kernel_bundle_impl>>()(
753  cl::sycl::detail::getSyclObjImpl(KernelBundle));
754  }
755 };
756 } // namespace std
cl::sycl::backend
backend
Definition: backend_types.hpp:21
cl::sycl::kernel_bundle::get_kernel
kernel get_kernel(const kernel_id &KernelID) const
Definition: kernel_bundle.hpp:255
cl::sycl::detail::get_kernel_bundle_impl
detail::KernelBundleImplPtr get_kernel_bundle_impl(const context &Ctx, const std::vector< device > &Devs, bundle_state State, const DevImgSelectorImpl &Selector)
Definition: kernel_bundle.cpp:135
cl::sycl::kernel_bundle::native_specialization_constant
bool native_specialization_constant() const noexcept
Definition: kernel_bundle.hpp:247
pi.h
cl::sycl::kernel_bundle
The kernel_bundle class represents collection of device images in a particular state.
Definition: kernel.hpp:28
cl::sycl::kernel_bundle::get_kernel_ids
std::vector< kernel_id > get_kernel_ids() const
Definition: kernel_bundle.hpp:235
cl::sycl::is_compatible
bool is_compatible(const device &Dev)
Definition: kernel_bundle.hpp:581
cl::sycl::kernel_bundle::get_context
context get_context() const noexcept
Definition: kernel_bundle.hpp:212
cl::sycl::distance
float distance(T p0, T p1) __NOEXC
Definition: builtins.hpp:1002
cl::sycl::backend_return_t
typename backend_traits< Backend >::template return_type< SyclType > backend_return_t
Definition: backend.hpp:65
cl::sycl::info::device
device
Definition: info_desc.hpp:50
cl::sycl::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:142
cl::sycl::kernel_bundle::has_kernel
bool has_kernel(const kernel_id &KernelID, const device &Dev) const noexcept
Definition: kernel_bundle.hpp:230
cl::sycl::detail::createSyclObjFromImpl
T createSyclObjFromImpl(decltype(T::impl) ImplObj)
Definition: common.hpp:215
device.hpp
cl::sycl::kernel_bundle::begin
device_image_iterator begin() const
Definition: kernel_bundle.hpp:303
__SYCL_DEPRECATED
#define __SYCL_DEPRECATED(message)
Definition: defines_elementary.hpp:47
cl::sycl::detail::device_image_plain
Definition: kernel_bundle.hpp:71
cl::sycl::join
sycl::kernel_bundle< State > join(const std::vector< sycl::kernel_bundle< State >> &Bundles)
Definition: kernel_bundle.hpp:604
context.hpp
cl::sycl::build
kernel_bundle< bundle_state::executable > build(const kernel_bundle< bundle_state::input > &InputBundle, const property_list &PropList={})
Definition: kernel_bundle.hpp:725
cl::sycl::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:245
detail
Definition: pi_opencl.cpp:86
cl::sycl::detail::KernelBundleImplPtr
std::shared_ptr< detail::kernel_bundle_impl > KernelBundleImplPtr
Definition: kernel_bundle.hpp:132
cl::sycl::device_image::has_kernel
bool has_kernel(const kernel_id &KernelID) const noexcept
Definition: kernel_bundle.hpp:109
cl::sycl::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:238
cl::sycl::property_list
Objects of the property_list class are containers for the SYCL properties.
Definition: property_list.hpp:26
pi.hpp
cl::sycl::detail::has_kernel_bundle_impl
bool has_kernel_bundle_impl(const context &Ctx, const std::vector< device > &Devs, const std::vector< kernel_id > &kernelIds, bundle_state State)
Definition: kernel_bundle.cpp:185
cast
To cast(From value)
Definition: pi_opencl.cpp:44
cl::sycl::detail::remove_reference_t
typename std::remove_reference< T >::type remove_reference_t
Definition: stl_type_traits.hpp:35
cl::sycl::bundle_state
bundle_state
Definition: kernel_bundle_enums.hpp:14
cl::sycl::detail::removeDuplicateDevices
const std::vector< device > removeDuplicateDevices(const std::vector< device > &Devs)
Definition: kernel_bundle.hpp:385
cl::sycl::kernel_bundle::contains_specialization_constants
bool contains_specialization_constants() const noexcept
Definition: kernel_bundle.hpp:241
cl::sycl::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:154
cl::sycl::kernel_bundle::has_kernel
bool has_kernel(const kernel_id &KernelID) const noexcept
Definition: kernel_bundle.hpp:223
kernel.hpp
cl::sycl::device
The SYCL device class encapsulates a single SYCL device on which kernels may be executed.
Definition: device.hpp:35
cl::sycl::device_image
Objects of the class represents an instance of an image in a specific state.
Definition: kernel_bundle.hpp:103
cl::sycl::kernel_id
Objects of the class identify kernel is some kernel_bundle related APIs.
Definition: kernel_bundle.hpp:39
cl::sycl::get_native
auto get_native(const SyclT &Obj) -> backend_return_t< Backend, SyclT >
kernel_bundle_enums.hpp
cl::sycl::detail::device_image_plain::operator==
bool operator==(const device_image_plain &RHS) const
Definition: kernel_bundle.hpp:76
cl::sycl::detail::device_image_plain::impl
detail::DeviceImageImplPtr impl
Definition: kernel_bundle.hpp:91
cl
We provide new interfaces for matrix muliply in this patch:
Definition: access.hpp:13
cl::sycl::detail::get_empty_interop_kernel_bundle
kernel_bundle< State > get_empty_interop_kernel_bundle(const context &Ctx)
make_kernel may need an empty interop kernel bundle.
Definition: kernel_bundle.hpp:484
cl::sycl::detail::kernel_bundle_plain::operator!=
bool operator!=(const kernel_bundle_plain &RHS) const
Definition: kernel_bundle.hpp:144
cl::sycl::detail::device_image_plain::operator!=
bool operator!=(const device_image_plain &RHS) const
Definition: kernel_bundle.hpp:80
cl::sycl::detail::kernel_bundle_plain::operator==
bool operator==(const kernel_bundle_plain &RHS) const
Definition: kernel_bundle.hpp:140
cl::sycl::detail::KernelInfo
Definition: kernel_desc.hpp:70
cl::sycl::get_kernel_ids
std::vector< kernel_id > get_kernel_ids()
Definition: kernel_bundle.cpp:286
pi_native_handle
uintptr_t pi_native_handle
Definition: pi.h:73
cl::sycl::context::get_devices
std::vector< device > get_devices() const
Gets devices associated with this SYCL context.
Definition: context.cpp:127
cl::sycl::has_kernel_bundle
bool has_kernel_bundle(const context &Ctx, const std::vector< device > &Devs)
Definition: kernel_bundle.hpp:569
cl::sycl::image_channel_order::a
@ a
std::hash< cl::sycl::device_image< State > >::operator()
size_t operator()(const cl::sycl::device_image< State > &DeviceImage) const
Definition: kernel_bundle.hpp:743
cl::sycl::get_kernel_bundle
kernel_bundle< State > get_kernel_bundle(const context &Ctx, SelectorT Selector)
Definition: kernel_bundle.hpp:512
cl::sycl::link
kernel_bundle< bundle_state::executable > link(const kernel_bundle< bundle_state::object > &ObjectBundle, const std::vector< device > &Devs, const property_list &PropList={})
Definition: kernel_bundle.hpp:693
cl::sycl::detail::DevImgSelectorImpl
std::function< bool(const detail::DeviceImageImplPtr &DevImgImpl)> DevImgSelectorImpl
Definition: kernel_bundle.hpp:467
cl::sycl::device_image::has_kernel
bool has_kernel(const kernel_id &KernelID, const device &Dev) const noexcept
Definition: kernel_bundle.hpp:115
cl::sycl::detail::getSyclObjImpl
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:198
cl::sycl::detail::find_device_intersection
std::vector< sycl::device > find_device_intersection(const std::vector< kernel_bundle< bundle_state::object >> &ObjectBundles)
Definition: kernel_bundle.cpp:260
std
Definition: accessor.hpp:2532
kernel_desc.hpp
cl::sycl::detail::kernel_bundle_plain::impl
detail::KernelBundleImplPtr impl
Definition: kernel_bundle.hpp:187
cl::sycl::kernel_bundle::empty
bool empty() const noexcept
Definition: kernel_bundle.hpp:204
cl::sycl::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:252
cl::sycl::kernel_bundle::get_backend
backend get_backend() const noexcept
Definition: kernel_bundle.hpp:207
cl::sycl::get_kernel_id
kernel_id get_kernel_id()
Definition: kernel_bundle.hpp:360
cl::sycl::context
The context class represents a SYCL context on which kernel functions may be executed.
Definition: context.hpp:35
std::hash< cl::sycl::kernel_id >::operator()
size_t operator()(const cl::sycl::kernel_id &KernelID) const
Definition: kernel_bundle.hpp:735
cl::sycl::kernel_bundle::get_devices
std::vector< device > get_devices() const noexcept
Definition: kernel_bundle.hpp:217
common.hpp
cl::sycl::kernel_id::operator!=
bool operator!=(const kernel_id &RHS) const
Definition: kernel_bundle.hpp:48
cl::sycl::detail::kernel_bundle_plain::kernel_bundle_plain
kernel_bundle_plain(const detail::KernelBundleImplPtr &Impl)
Definition: kernel_bundle.hpp:137
cl::sycl::info::context
context
Definition: info_desc.hpp:41
std::hash< cl::sycl::kernel_bundle< State > >::operator()
size_t operator()(const cl::sycl::kernel_bundle< State > &KernelBundle) const
Definition: kernel_bundle.hpp:751
cl::sycl::detail::kernel_bundle_plain
Definition: kernel_bundle.hpp:135
cl::sycl::detail::enable_if_t
typename std::enable_if< B, T >::type enable_if_t
Definition: stl_type_traits.hpp:24
cl::sycl::detail::device_image_plain::device_image_plain
device_image_plain(const detail::DeviceImageImplPtr &Impl)
Definition: kernel_bundle.hpp:73
cl::sycl::kernel
Provides an abstraction of a SYCL kernel.
Definition: kernel.hpp:67
cl::sycl::kernel_bundle::end
device_image_iterator end() const
Definition: kernel_bundle.hpp:309
cl::sycl::detail::get_kernel_id_impl
kernel_id get_kernel_id_impl(std::string KernelName)
Definition: kernel_bundle.cpp:116
cl::sycl::detail::getDeviceComparisonLambda
auto getDeviceComparisonLambda()
Definition: kernel_bundle.hpp:380
cl::sycl::kernel_id::operator==
bool operator==(const kernel_id &RHS) const
Definition: kernel_bundle.hpp:46
cl::sycl::compile
kernel_bundle< bundle_state::object > compile(const kernel_bundle< bundle_state::input > &InputBundle, const property_list &PropList={})
Definition: kernel_bundle.hpp:643
__SYCL_INLINE_NAMESPACE
#define __SYCL_INLINE_NAMESPACE(X)
Definition: defines_elementary.hpp:12
cl::sycl::detail::DeviceImageImplPtr
std::shared_ptr< device_image_impl > DeviceImageImplPtr
Definition: kernel_bundle.hpp:67