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