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 
11 
12 #include <set>
13 
15 namespace sycl {
16 
17 kernel_id::kernel_id(const char *Name)
18  : impl(std::make_shared<detail::kernel_id_impl>(Name)) {}
19 
20 const char *kernel_id::get_name() const noexcept { return impl->get_name(); }
21 
22 namespace detail {
23 
27 
28 bool device_image_plain::has_kernel(const kernel_id &KernelID) const noexcept {
29  return impl->has_kernel(KernelID);
30 }
31 
32 bool device_image_plain::has_kernel(const kernel_id &KernelID,
33  const device &Dev) const noexcept {
34  return impl->has_kernel(KernelID, Dev);
35 }
36 
37 pi_native_handle device_image_plain::getNative() const {
38  return impl->getNative();
39 }
40 
44 
45 bool kernel_bundle_plain::empty() const noexcept { return impl->empty(); }
46 
47 backend kernel_bundle_plain::get_backend() const noexcept {
48  return impl->get_backend();
49 }
50 
51 context kernel_bundle_plain::get_context() const noexcept {
52  return impl->get_context();
53 }
54 
55 std::vector<device> kernel_bundle_plain::get_devices() const noexcept {
56  return impl->get_devices();
57 }
58 
59 std::vector<kernel_id> kernel_bundle_plain::get_kernel_ids() const {
60  return impl->get_kernel_ids();
61 }
62 
63 bool kernel_bundle_plain::contains_specialization_constants() const noexcept {
64  return impl->contains_specialization_constants();
65 }
66 
67 bool kernel_bundle_plain::native_specialization_constant() const noexcept {
68  return impl->native_specialization_constant();
69 }
70 
71 kernel kernel_bundle_plain::get_kernel(const kernel_id &KernelID) const {
72  return impl->get_kernel(KernelID, impl);
73 }
74 
75 const device_image_plain *kernel_bundle_plain::begin() const {
76  return impl->begin();
77 }
78 
79 const device_image_plain *kernel_bundle_plain::end() const {
80  return impl->end();
81 }
82 
83 bool kernel_bundle_plain::has_kernel(const kernel_id &KernelID) const noexcept {
84  return impl->has_kernel(KernelID);
85 }
86 
87 bool kernel_bundle_plain::has_kernel(const kernel_id &KernelID,
88  const device &Dev) const noexcept {
89  return impl->has_kernel(KernelID, Dev);
90 }
91 
92 bool kernel_bundle_plain::has_specialization_constant_impl(
93  const char *SpecName) const noexcept {
94  return impl->has_specialization_constant(SpecName);
95 }
96 
97 void kernel_bundle_plain::set_specialization_constant_impl(
98  const char *SpecName, void *Value, size_t Size) noexcept {
99  impl->set_specialization_constant_raw_value(SpecName, Value, Size);
100 }
101 
102 void kernel_bundle_plain::get_specialization_constant_impl(
103  const char *SpecName, void *Value) const noexcept {
104  impl->get_specialization_constant_raw_value(SpecName, Value);
105 }
106 
107 bool kernel_bundle_plain::is_specialization_constant_set(
108  const char *SpecName) const noexcept {
109  return impl->is_specialization_constant_set(SpecName);
110 }
111 
115 
116 kernel_id get_kernel_id_impl(std::string KernelName) {
117  return detail::ProgramManager::getInstance().getSYCLKernelID(KernelName);
118 }
119 
121 get_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
122  bundle_state State) {
123  return std::make_shared<detail::kernel_bundle_impl>(Ctx, Devs, State);
124 }
125 
127 get_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
128  const std::vector<kernel_id> &KernelIDs,
129  bundle_state State) {
130  return std::make_shared<detail::kernel_bundle_impl>(Ctx, Devs, KernelIDs,
131  State);
132 }
133 
135 get_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
136  bundle_state State, const DevImgSelectorImpl &Selector) {
137  return std::make_shared<detail::kernel_bundle_impl>(Ctx, Devs, Selector,
138  State);
139 }
140 
143  const std::vector<device> &Devs) {
144  return std::make_shared<detail::kernel_bundle_impl>(Ctx, Devs);
145 }
146 
147 std::shared_ptr<detail::kernel_bundle_impl>
148 join_impl(const std::vector<detail::KernelBundleImplPtr> &Bundles) {
149  return std::make_shared<detail::kernel_bundle_impl>(Bundles,
150  bundle_state::input);
151 }
152 
153 std::shared_ptr<detail::kernel_bundle_impl>
154 join_impl(const std::vector<detail::KernelBundleImplPtr> &Bundles,
155  bundle_state State) {
156  return std::make_shared<detail::kernel_bundle_impl>(Bundles, State);
157 }
158 
159 bool has_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
160  bundle_state State) {
161  // Check that all requested devices are associated with the context
162  const bool AllDevicesInTheContext = checkAllDevicesAreInContext(Devs, Ctx);
163  if (Devs.empty() || !AllDevicesInTheContext)
164  throw sycl::exception(make_error_code(errc::invalid),
165  "Not all devices are associated with the context or "
166  "vector of devices is empty");
167 
168  if (bundle_state::input == State &&
169  !checkAllDevicesHaveAspect(Devs, aspect::online_compiler))
170  return false;
171  if (bundle_state::object == State &&
172  !checkAllDevicesHaveAspect(Devs, aspect::online_linker))
173  return false;
174 
175  const std::vector<device_image_plain> DeviceImages =
176  detail::ProgramManager::getInstance()
177  .getSYCLDeviceImagesWithCompatibleState(Ctx, Devs, State);
178 
179  // TODO: Add a check that all kernel ids are compatible with at least one
180  // device in Devs
181 
182  return (bool)DeviceImages.size();
183 }
184 
185 bool has_kernel_bundle_impl(const context &Ctx, const std::vector<device> &Devs,
186  const std::vector<kernel_id> &KernelIds,
187  bundle_state State) {
188  // Check that all requested devices are associated with the context
189  const bool AllDevicesInTheContext = checkAllDevicesAreInContext(Devs, Ctx);
190 
191  if (Devs.empty() || !AllDevicesInTheContext)
192  throw sycl::exception(make_error_code(errc::invalid),
193  "Not all devices are associated with the context or "
194  "vector of devices is empty");
195 
196  bool DeviceHasRequireAspectForState = true;
197  if (bundle_state::input == State) {
198  DeviceHasRequireAspectForState =
199  std::all_of(Devs.begin(), Devs.end(), [](const device &Dev) {
200  return Dev.has(aspect::online_compiler);
201  });
202  } else if (bundle_state::object == State) {
203  DeviceHasRequireAspectForState =
204  std::all_of(Devs.begin(), Devs.end(), [](const device &Dev) {
205  return Dev.has(aspect::online_linker);
206  });
207  }
208 
209  if (!DeviceHasRequireAspectForState)
210  return false;
211 
212  const std::vector<device_image_plain> DeviceImages =
213  detail::ProgramManager::getInstance()
214  .getSYCLDeviceImagesWithCompatibleState(Ctx, Devs, State);
215 
216  std::set<kernel_id, LessByNameComp> CombinedKernelIDs;
217  for (const device_image_plain &DeviceImage : DeviceImages) {
218  const std::shared_ptr<device_image_impl> &DeviceImageImpl =
219  getSyclObjImpl(DeviceImage);
220 
221  CombinedKernelIDs.insert(DeviceImageImpl->get_kernel_ids_ptr()->begin(),
222  DeviceImageImpl->get_kernel_ids_ptr()->end());
223  }
224 
225  const bool AllKernelIDsRepresented =
226  std::all_of(KernelIds.begin(), KernelIds.end(),
227  [&CombinedKernelIDs](const kernel_id &KernelID) {
228  return CombinedKernelIDs.count(KernelID);
229  });
230 
231  // TODO: Add a check that all kernel ids are compatible with at least one
232  // device in Devs
233 
234  return AllKernelIDsRepresented;
235 }
236 
237 std::shared_ptr<detail::kernel_bundle_impl>
239  const std::vector<device> &Devs, const property_list &PropList) {
240  return std::make_shared<detail::kernel_bundle_impl>(
241  InputBundle, Devs, PropList, bundle_state::object);
242 }
243 
244 std::shared_ptr<detail::kernel_bundle_impl>
245 link_impl(const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles,
246  const std::vector<device> &Devs, const property_list &PropList) {
247  return std::make_shared<detail::kernel_bundle_impl>(ObjectBundles, Devs,
248  PropList);
249 }
250 
251 std::shared_ptr<detail::kernel_bundle_impl>
253  const std::vector<device> &Devs, const property_list &PropList) {
254  return std::make_shared<detail::kernel_bundle_impl>(
255  InputBundle, Devs, PropList, bundle_state::executable);
256 }
257 
258 // This function finds intersection of associated devices in common for all
259 // bundles
260 std::vector<sycl::device> find_device_intersection(
261  const std::vector<kernel_bundle<bundle_state::object>> &ObjectBundles) {
262  std::vector<sycl::device> IntersectDevices;
263  std::vector<unsigned int> DevsCounters;
264  std::map<device, unsigned int, LessByHash<device>> DevCounters;
265  for (const sycl::kernel_bundle<bundle_state::object> &ObjectBundle :
266  ObjectBundles)
267  // Increment counter in "DevCounters" each time a device is seen
268  for (const sycl::device &Device : ObjectBundle.get_devices())
269  DevCounters[Device]++;
270 
271  // If some device counter is less than ObjectBundles.size() then some bundle
272  // doesn't have it - do not add such a device to the final result
273  for (const std::pair<const device, unsigned int> &It : DevCounters)
274  if (ObjectBundles.size() == It.second)
275  IntersectDevices.push_back(It.first);
276 
277  return IntersectDevices;
278 }
279 
280 } // namespace detail
281 
285 
286 std::vector<kernel_id> get_kernel_ids() {
287  return detail::ProgramManager::getInstance().getAllSYCLKernelIDs();
288 }
289 
290 } // namespace sycl
291 } // __SYCL_INLINE_NAMESPACE(cl)
cl::sycl::backend
backend
Definition: backend_types.hpp:21
cl::sycl::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:121
cl::sycl::kernel_bundle
The kernel_bundle class represents collection of device images in a particular state.
Definition: kernel.hpp:29
cl::sycl::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:159
cl::sycl::detail::device_image_plain::has_kernel
bool has_kernel(const kernel_id &KernelID) const noexcept
Definition: kernel_bundle.cpp:28
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::detail::device_image_plain
Definition: kernel_bundle.hpp:71
cl::sycl::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:121
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
cl::sycl::detail::KernelBundleImplPtr
std::shared_ptr< detail::kernel_bundle_impl > KernelBundleImplPtr
Definition: kernel_bundle.hpp:132
sycl
Definition: invoke_simd.hpp:68
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::detail::checkAllDevicesHaveAspect
static bool checkAllDevicesHaveAspect(const std::vector< device > &Devices, aspect Aspect)
Definition: kernel_bundle_impl.hpp:41
kernel_id_impl.hpp
cl::sycl::property_list
Objects of the property_list class are containers for the SYCL properties.
Definition: property_list.hpp:26
cl::sycl::bundle_state
bundle_state
Definition: kernel_bundle_enums.hpp:14
cl::sycl::detail::checkAllDevicesAreInContext
static bool checkAllDevicesAreInContext(const std::vector< device > &Devices, const context &Context)
Definition: kernel_bundle_impl.hpp:31
cl::sycl::device
The SYCL device class encapsulates a single SYCL device on which kernels may be executed.
Definition: device.hpp:36
cl::sycl::kernel_id
Objects of the class identify kernel is some kernel_bundle related APIs.
Definition: kernel_bundle.hpp:39
Device
@ Device
Definition: usm_allocator.hpp:14
cl
We provide new interfaces for matrix muliply in this patch:
Definition: access.hpp:13
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:98
cl::sycl::context::get_devices
std::vector< device > get_devices() const
Gets devices associated with this SYCL context.
Definition: context.cpp:128
all_of
bool all_of(const simd_mask< _Tp, _Abi > &) noexcept
cl::sycl::detail::DevImgSelectorImpl
std::function< bool(const detail::DeviceImageImplPtr &DevImgImpl)> DevImgSelectorImpl
Definition: kernel_bundle.hpp:467
cl::sycl::detail::getSyclObjImpl
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:204
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:2617
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::context
The context class represents a SYCL context on which kernel functions may be executed.
Definition: context.hpp:38
kernel_bundle_impl.hpp
cl::sycl::exception
Definition: exception.hpp:64
cl::sycl::detail::join_impl
std::shared_ptr< detail::kernel_bundle_impl > join_impl(const std::vector< detail::KernelBundleImplPtr > &Bundles)
Definition: kernel_bundle.cpp:148
cl::sycl::kernel
Provides an abstraction of a SYCL kernel.
Definition: kernel.hpp:68
cl::sycl::detail::get_kernel_id_impl
kernel_id get_kernel_id_impl(std::string KernelName)
Definition: kernel_bundle.cpp:116
__SYCL_INLINE_NAMESPACE
#define __SYCL_INLINE_NAMESPACE(X)
Definition: defines_elementary.hpp:12