alpaka
Abstraction Library for Parallel Kernel Acceleration
PlatformGenericSycl.hpp
Go to the documentation of this file.
1 /* Copyright 2024 Jan Stephan, Luca Ferragina, Aurora Perego
2  * SPDX-License-Identifier: MPL-2.0
3  */
4 
5 #pragma once
6 
7 #include "alpaka/core/Interface.hpp"
8 #include "alpaka/core/Sycl.hpp"
9 #include "alpaka/dev/DevGenericSycl.hpp"
10 #include "alpaka/dev/Traits.hpp"
11 #include "alpaka/platform/Traits.hpp"
12 
13 #include <cstddef>
14 #include <exception>
15 #if ALPAKA_DEBUG >= ALPAKA_DEBUG_MINIMAL
16 # include <iostream>
17 #endif
18 #include <sstream>
19 #include <stdexcept>
20 #include <vector>
21 
22 #ifdef ALPAKA_ACC_SYCL_ENABLED
23 
24 # if BOOST_COMP_CLANG
25 # pragma clang diagnostic push
26 # pragma clang diagnostic ignored "-Wswitch-default"
27 # endif
28 
29 # include <sycl/sycl.hpp>
30 
31 namespace alpaka
32 {
33  namespace detail
34  {
35  template<concepts::Tag TTag>
36  struct SYCLDeviceSelector;
37  } // namespace detail
38 
39  //! The SYCL device manager.
40  template<concepts::Tag TTag>
41  struct PlatformGenericSycl : interface::Implements<ConceptPlatform, PlatformGenericSycl<TTag>>
42  {
43  PlatformGenericSycl()
44  : platform{detail::SYCLDeviceSelector<TTag>{}}
45  , devices(platform.get_devices())
46  , context{sycl::context{
47  devices,
48  [](sycl::exception_list exceptions)
49  {
50  auto ss_err = std::stringstream{};
51  ss_err << "Caught asynchronous SYCL exception(s):\n";
52  for(std::exception_ptr e : exceptions)
53  {
54  try
55  {
56  std::rethrow_exception(e);
57  }
58  catch(sycl::exception const& err)
59  {
60  ss_err << err.what() << " (" << err.code() << ")\n";
61  }
62  }
63  throw std::runtime_error(ss_err.str());
64  }}}
65  {
66  }
67 
68  [[nodiscard]] auto syclPlatform() -> sycl::platform&
69  {
70  return platform;
71  }
72 
73  [[nodiscard]] auto syclPlatform() const -> sycl::platform const&
74  {
75  return platform;
76  }
77 
78  [[nodiscard]] auto syclDevices() -> std::vector<sycl::device>&
79  {
80  return devices;
81  }
82 
83  [[nodiscard]] auto syclDevices() const -> std::vector<sycl::device> const&
84  {
85  return devices;
86  }
87 
88  [[nodiscard]] auto syclContext() -> sycl::context&
89  {
90  return context;
91  }
92 
93  [[nodiscard]] auto syclContext() const -> sycl::context const&
94  {
95  return context;
96  }
97 
98  private:
99  sycl::platform platform;
100  std::vector<sycl::device> devices;
101  sycl::context context;
102  };
103 
104  namespace trait
105  {
106  //! The SYCL platform device type trait specialization.
107  template<concepts::Tag TTag>
108  struct DevType<PlatformGenericSycl<TTag>>
109  {
110  using type = DevGenericSycl<TTag>;
111  };
112 
113  //! The SYCL platform device count get trait specialization.
114  template<concepts::Tag TTag>
115  struct GetDevCount<PlatformGenericSycl<TTag>>
116  {
117  static auto getDevCount(PlatformGenericSycl<TTag> const& platform) -> std::size_t
118  {
119  ALPAKA_DEBUG_FULL_LOG_SCOPE;
120 
121  return platform.syclDevices().size();
122  }
123  };
124 
125  //! The SYCL platform device get trait specialization.
126  template<concepts::Tag TTag>
127  struct GetDevByIdx<PlatformGenericSycl<TTag>>
128  {
129  static auto getDevByIdx(PlatformGenericSycl<TTag> const& platform, std::size_t const& devIdx)
130  {
131  ALPAKA_DEBUG_FULL_LOG_SCOPE;
132 
133  auto const& devices = platform.syclDevices();
134  if(devIdx >= devices.size())
135  {
136  auto ss_err = std::stringstream{};
137  ss_err << "Unable to return device handle for device " << devIdx << ". There are only "
138  << devices.size() << " SYCL devices!";
139  throw std::runtime_error(ss_err.str());
140  }
141 
142  auto sycl_dev = devices.at(devIdx);
143 
144  // Log this device.
145 # if ALPAKA_DEBUG >= ALPAKA_DEBUG_FULL
146  printDeviceProperties(sycl_dev);
147 # elif ALPAKA_DEBUG >= ALPAKA_DEBUG_MINIMAL
148  std::cout << __func__ << sycl_dev.template get_info<sycl::info::device::name>() << '\n';
149 # endif
150  using SyclPlatform = alpaka::PlatformGenericSycl<TTag>;
151  return typename DevType<SyclPlatform>::type{sycl_dev, platform.syclContext()};
152  }
153 
154  private:
155 # if ALPAKA_DEBUG >= ALPAKA_DEBUG_FULL
156  //! Prints all the device properties to std::cout.
157  static auto printDeviceProperties(sycl::device const& device) -> void
158  {
159  ALPAKA_DEBUG_FULL_LOG_SCOPE;
160 
161  constexpr auto KiB = std::size_t{1024};
162  constexpr auto MiB = KiB * KiB;
163 
164  std::cout << "Device type: ";
165  switch(device.get_info<sycl::info::device::device_type>())
166  {
167  case sycl::info::device_type::cpu:
168  std::cout << "CPU";
169  break;
170 
171  case sycl::info::device_type::gpu:
172  std::cout << "GPU";
173  break;
174 
175  case sycl::info::device_type::accelerator:
176  std::cout << "Accelerator";
177  break;
178 
179  case sycl::info::device_type::custom:
180  std::cout << "Custom";
181  break;
182 
183  case sycl::info::device_type::automatic:
184  std::cout << "Automatic";
185  break;
186 
187  case sycl::info::device_type::host:
188  std::cout << "Host";
189  break;
190 
191  // The SYCL spec forbids the return of device_type::all
192  // Including this here to prevent warnings because of
193  // missing cases
194  case sycl::info::device_type::all:
195  std::cout << "All";
196  break;
197  }
198  std::cout << '\n';
199 
200  std::cout << "Name: " << device.get_info<sycl::info::device::name>() << '\n';
201 
202  std::cout << "Vendor: " << device.get_info<sycl::info::device::vendor>() << '\n';
203 
204  std::cout << "Vendor ID: " << device.get_info<sycl::info::device::vendor_id>() << '\n';
205 
206  std::cout << "Driver version: " << device.get_info<sycl::info::device::driver_version>() << '\n';
207 
208  std::cout << "SYCL version: " << device.get_info<sycl::info::device::version>() << '\n';
209 
210 # if !defined(BOOST_COMP_ICPX)
211  // Not defined by Level Zero back-end
212  std::cout << "Backend version: " << device.get_info<sycl::info::device::backend_version>() << '\n';
213 # endif
214 
215  std::cout << "Aspects: " << '\n';
216 
217 # if defined(BOOST_COMP_ICPX)
218 # if BOOST_COMP_ICPX >= BOOST_VERSION_NUMBER(53, 2, 0)
219  // These aspects are missing from oneAPI versions < 2023.2.0
220  if(device.has(sycl::aspect::emulated))
221  std::cout << "\t* emulated\n";
222 
223  if(device.has(sycl::aspect::host_debuggable))
224  std::cout << "\t* debuggable using standard debuggers\n";
225 # endif
226 # endif
227 
228  if(device.has(sycl::aspect::fp16))
229  std::cout << "\t* supports sycl::half precision\n";
230 
231  if(device.has(sycl::aspect::fp64))
232  std::cout << "\t* supports double precision\n";
233 
234  if(device.has(sycl::aspect::atomic64))
235  std::cout << "\t* supports 64-bit atomics\n";
236 
237  if(device.has(sycl::aspect::image))
238  std::cout << "\t* supports images\n";
239 
240  if(device.has(sycl::aspect::online_compiler))
241  std::cout << "\t* supports online compilation of device code\n";
242 
243  if(device.has(sycl::aspect::online_linker))
244  std::cout << "\t* supports online linking of device code\n";
245 
246  if(device.has(sycl::aspect::queue_profiling))
247  std::cout << "\t* supports queue profiling\n";
248 
249  if(device.has(sycl::aspect::usm_device_allocations))
250  std::cout << "\t* supports explicit USM allocations\n";
251 
252  if(device.has(sycl::aspect::usm_host_allocations))
253  std::cout << "\t* can access USM memory allocated by sycl::usm::alloc::host\n";
254 
255  if(device.has(sycl::aspect::usm_atomic_host_allocations))
256  std::cout << "\t* can access USM memory allocated by sycl::usm::alloc::host atomically\n";
257 
258  if(device.has(sycl::aspect::usm_shared_allocations))
259  std::cout << "\t* can access USM memory allocated by sycl::usm::alloc::shared\n";
260 
261  if(device.has(sycl::aspect::usm_atomic_shared_allocations))
262  std::cout << "\t* can access USM memory allocated by sycl::usm::alloc::shared atomically\n";
263 
264  if(device.has(sycl::aspect::usm_system_allocations))
265  std::cout << "\t* can access memory allocated by the system allocator\n";
266 
267  std::cout << "Available compute units: " << device.get_info<sycl::info::device::max_compute_units>()
268  << '\n';
269 
270  std::cout << "Maximum work item dimensions: ";
271  auto dims = device.get_info<sycl::info::device::max_work_item_dimensions>();
272  std::cout << dims << std::endl;
273 
274  std::cout << "Maximum number of work items:\n";
275  auto const wi_1D = device.get_info<sycl::info::device::max_work_item_sizes<1>>();
276  auto const wi_2D = device.get_info<sycl::info::device::max_work_item_sizes<2>>();
277  auto const wi_3D = device.get_info<sycl::info::device::max_work_item_sizes<3>>();
278  std::cout << "\t* 1D: (" << wi_1D.get(0) << ")\n";
279  std::cout << "\t* 2D: (" << wi_2D.get(0) << ", " << wi_2D.get(1) << ")\n";
280  std::cout << "\t* 3D: (" << wi_3D.get(0) << ", " << wi_3D.get(1) << ", " << wi_3D.get(2) << ")\n";
281 
282  std::cout << "Maximum number of work items per work-group: "
283  << device.get_info<sycl::info::device::max_work_group_size>() << '\n';
284 
285  std::cout << "Maximum number of sub-groups per work-group: "
286  << device.get_info<sycl::info::device::max_num_sub_groups>() << '\n';
287 
288  std::cout << "Supported sub-group sizes: ";
289  auto const sg_sizes = device.get_info<sycl::info::device::sub_group_sizes>();
290  for(auto const& sz : sg_sizes)
291  std::cout << sz << ", ";
292  std::cout << '\n';
293 
294  std::cout << "Preferred native vector width (char): "
295  << device.get_info<sycl::info::device::preferred_vector_width_char>() << '\n';
296 
297  std::cout << "Native ISA vector width (char): "
298  << device.get_info<sycl::info::device::native_vector_width_char>() << '\n';
299 
300  std::cout << "Preferred native vector width (short): "
301  << device.get_info<sycl::info::device::preferred_vector_width_short>() << '\n';
302 
303  std::cout << "Native ISA vector width (short): "
304  << device.get_info<sycl::info::device::native_vector_width_short>() << '\n';
305 
306  std::cout << "Preferred native vector width (int): "
307  << device.get_info<sycl::info::device::preferred_vector_width_int>() << '\n';
308 
309  std::cout << "Native ISA vector width (int): "
310  << device.get_info<sycl::info::device::native_vector_width_int>() << '\n';
311 
312  std::cout << "Preferred native vector width (long): "
313  << device.get_info<sycl::info::device::preferred_vector_width_long>() << '\n';
314 
315  std::cout << "Native ISA vector width (long): "
316  << device.get_info<sycl::info::device::native_vector_width_long>() << '\n';
317 
318  std::cout << "Preferred native vector width (float): "
319  << device.get_info<sycl::info::device::preferred_vector_width_float>() << '\n';
320 
321  std::cout << "Native ISA vector width (float): "
322  << device.get_info<sycl::info::device::native_vector_width_float>() << '\n';
323 
324  if(device.has(sycl::aspect::fp64))
325  {
326  std::cout << "Preferred native vector width (double): "
327  << device.get_info<sycl::info::device::preferred_vector_width_double>() << '\n';
328 
329  std::cout << "Native ISA vector width (double): "
330  << device.get_info<sycl::info::device::native_vector_width_double>() << '\n';
331  }
332 
333  if(device.has(sycl::aspect::fp16))
334  {
335  std::cout << "Preferred native vector width (half): "
336  << device.get_info<sycl::info::device::preferred_vector_width_half>() << '\n';
337 
338  std::cout << "Native ISA vector width (half): "
339  << device.get_info<sycl::info::device::native_vector_width_half>() << '\n';
340  }
341 
342  std::cout << "Maximum clock frequency: " << device.get_info<sycl::info::device::max_clock_frequency>()
343  << " MHz\n";
344 
345  std::cout << "Address space size: " << device.get_info<sycl::info::device::address_bits>() << "-bit\n";
346 
347  std::cout << "Maximum size of memory object allocation: "
348  << device.get_info<sycl::info::device::max_mem_alloc_size>() << " bytes\n";
349 
350  if(device.has(sycl::aspect::image))
351  {
352  std::cout << "Maximum number of simultaneous image object reads per kernel: "
353  << device.get_info<sycl::info::device::max_read_image_args>() << '\n';
354 
355  std::cout << "Maximum number of simultaneous image writes per kernel: "
356  << device.get_info<sycl::info::device::max_write_image_args>() << '\n';
357 
358  std::cout << "Maximum 1D/2D image width: "
359  << device.get_info<sycl::info::device::image2d_max_width>() << " px\n";
360 
361  std::cout << "Maximum 2D image height: "
362  << device.get_info<sycl::info::device::image2d_max_height>() << " px\n";
363 
364  std::cout << "Maximum 3D image width: " << device.get_info<sycl::info::device::image3d_max_width>()
365  << " px\n";
366 
367  std::cout << "Maximum 3D image height: "
368  << device.get_info<sycl::info::device::image3d_max_height>() << " px\n";
369 
370  std::cout << "Maximum 3D image depth: " << device.get_info<sycl::info::device::image3d_max_depth>()
371  << " px\n";
372 
373  std::cout << "Maximum number of samplers per kernel: "
374  << device.get_info<sycl::info::device::max_samplers>() << '\n';
375  }
376 
377  std::cout << "Maximum kernel argument size: "
378  << device.get_info<sycl::info::device::max_parameter_size>() << " bytes\n";
379 
380  std::cout << "Memory base address alignment: "
381  << device.get_info<sycl::info::device::mem_base_addr_align>() << " bit\n";
382 
383  auto print_fp_config = [](std::string const& fp, std::vector<sycl::info::fp_config> const& conf)
384  {
385  std::cout << fp << " precision floating-point capabilities:\n";
386 
387  auto find_and_print = [&](sycl::info::fp_config val)
388  {
389  auto it = std::find(begin(conf), end(conf), val);
390  std::cout << (it == std::end(conf) ? "No" : "Yes") << '\n';
391  };
392 
393  std::cout << "\t* denorm support: ";
394  find_and_print(sycl::info::fp_config::denorm);
395 
396  std::cout << "\t* INF & quiet NaN support: ";
397  find_and_print(sycl::info::fp_config::inf_nan);
398 
399  std::cout << "\t* round to nearest even support: ";
400  find_and_print(sycl::info::fp_config::round_to_nearest);
401 
402  std::cout << "\t* round to zero support: ";
403  find_and_print(sycl::info::fp_config::round_to_zero);
404 
405  std::cout << "\t* round to infinity support: ";
406  find_and_print(sycl::info::fp_config::round_to_inf);
407 
408  std::cout << "\t* IEEE754-2008 FMA support: ";
409  find_and_print(sycl::info::fp_config::fma);
410 
411  std::cout << "\t* correctly rounded divide/sqrt support: ";
412  find_and_print(sycl::info::fp_config::correctly_rounded_divide_sqrt);
413 
414  std::cout << "\t* software-implemented floating point operations: ";
415  find_and_print(sycl::info::fp_config::soft_float);
416  };
417 
418  if(device.has(sycl::aspect::fp16))
419  {
420  auto const fp16_conf = device.get_info<sycl::info::device::half_fp_config>();
421  print_fp_config("Half", fp16_conf);
422  }
423 
424  auto const fp32_conf = device.get_info<sycl::info::device::single_fp_config>();
425  print_fp_config("Single", fp32_conf);
426 
427  if(device.has(sycl::aspect::fp64))
428  {
429  auto const fp64_conf = device.get_info<sycl::info::device::double_fp_config>();
430  print_fp_config("Double", fp64_conf);
431  }
432 
433  std::cout << "Global memory cache type: ";
434  auto has_global_mem_cache = false;
435  switch(device.get_info<sycl::info::device::global_mem_cache_type>())
436  {
437  case sycl::info::global_mem_cache_type::none:
438  std::cout << "none";
439  break;
440 
441  case sycl::info::global_mem_cache_type::read_only:
442  std::cout << "read-only";
443  has_global_mem_cache = true;
444  break;
445 
446  case sycl::info::global_mem_cache_type::read_write:
447  std::cout << "read-write";
448  has_global_mem_cache = true;
449  break;
450  }
451  std::cout << '\n';
452 
453  if(has_global_mem_cache)
454  {
455  std::cout << "Global memory cache line size: "
456  << device.get_info<sycl::info::device::global_mem_cache_line_size>() << " bytes\n";
457 
458  std::cout << "Global memory cache size: "
459  << device.get_info<sycl::info::device::global_mem_cache_size>() / KiB << " KiB\n";
460  }
461 
462  std::cout << "Global memory size: " << device.get_info<sycl::info::device::global_mem_size>() / MiB
463  << " MiB" << std::endl;
464 
465  std::cout << "Local memory type: ";
466  auto has_local_memory = false;
467  switch(device.get_info<sycl::info::device::local_mem_type>())
468  {
469  case sycl::info::local_mem_type::none:
470  std::cout << "none";
471  break;
472 
473  case sycl::info::local_mem_type::local:
474  std::cout << "local";
475  has_local_memory = true;
476  break;
477 
478  case sycl::info::local_mem_type::global:
479  std::cout << "global";
480  has_local_memory = true;
481  break;
482  }
483  std::cout << '\n';
484 
485  if(has_local_memory)
486  std::cout << "Local memory size: " << device.get_info<sycl::info::device::local_mem_size>() / KiB
487  << " KiB\n";
488 
489  std::cout << "Error correction support: "
490  << (device.get_info<sycl::info::device::error_correction_support>() ? "Yes" : "No") << '\n';
491 
492  auto print_memory_orders = [](std::vector<sycl::memory_order> const& mem_orders)
493  {
494  for(auto const& cap : mem_orders)
495  {
496  switch(cap)
497  {
498  case sycl::memory_order::relaxed:
499  std::cout << "relaxed";
500  break;
501 
502  case sycl::memory_order::acquire:
503  std::cout << "acquire";
504  break;
505 
506  case sycl::memory_order::release:
507  std::cout << "release";
508  break;
509 
510  case sycl::memory_order::acq_rel:
511  std::cout << "acq_rel";
512  break;
513 
514  case sycl::memory_order::seq_cst:
515  std::cout << "seq_cst";
516  break;
517 # if defined(BOOST_COMP_ICPX)
518  // Stop icpx from complaining about its own internals.
519  case sycl::memory_order::__consume_unsupported:
520  break;
521 # endif
522  }
523  std::cout << ", ";
524  }
525  std::cout << '\n';
526  };
527 
528  std::cout << "Supported memory orderings for atomic operations: ";
529  auto const mem_orders = device.get_info<sycl::info::device::atomic_memory_order_capabilities>();
530  print_memory_orders(mem_orders);
531 
532 # if defined(BOOST_COMP_ICPX)
533 # if BOOST_COMP_ICPX >= BOOST_VERSION_NUMBER(53, 2, 0)
534  // Not implemented in oneAPI < 2023.2.0
535  std::cout << "Supported memory orderings for sycl::atomic_fence: ";
536  auto const fence_orders = device.get_info<sycl::info::device::atomic_fence_order_capabilities>();
537  print_memory_orders(fence_orders);
538 # endif
539 # endif
540 
541  auto print_memory_scopes = [](std::vector<sycl::memory_scope> const& mem_scopes)
542  {
543  for(auto const& cap : mem_scopes)
544  {
545  switch(cap)
546  {
547  case sycl::memory_scope::work_item:
548  std::cout << "work-item";
549  break;
550 
551  case sycl::memory_scope::sub_group:
552  std::cout << "sub-group";
553  break;
554 
555  case sycl::memory_scope::work_group:
556  std::cout << "work-group";
557  break;
558 
559  case sycl::memory_scope::device:
560  std::cout << "device";
561  break;
562 
563  case sycl::memory_scope::system:
564  std::cout << "system";
565  break;
566  }
567  std::cout << ", ";
568  }
569  std::cout << '\n';
570  };
571 
572  std::cout << "Supported memory scopes for atomic operations: ";
573  auto const mem_scopes = device.get_info<sycl::info::device::atomic_memory_scope_capabilities>();
574  print_memory_scopes(mem_scopes);
575 
576 # if defined(BOOST_COMP_ICPX)
577 # if BOOST_COMP_ICPX >= BOOST_VERSION_NUMBER(53, 2, 0)
578  // Not implemented in oneAPI < 2023.2.0
579  std::cout << "Supported memory scopes for sycl::atomic_fence: ";
580  auto const fence_scopes = device.get_info<sycl::info::device::atomic_fence_scope_capabilities>();
581  print_memory_scopes(fence_scopes);
582 # endif
583 # endif
584 
585  std::cout << "Device timer resolution: "
586  << device.get_info<sycl::info::device::profiling_timer_resolution>() << " ns\n";
587 
588  std::cout << "Built-in kernels: ";
589  auto const builtins = device.get_info<sycl::info::device::built_in_kernel_ids>();
590  for(auto const& b : builtins)
591  std::cout << b.get_name() << ", ";
592  std::cout << '\n';
593 
594  std::cout << "Maximum number of subdevices: ";
595  auto const max_subs = device.get_info<sycl::info::device::partition_max_sub_devices>();
596  std::cout << max_subs << '\n';
597 
598  if(max_subs > 1)
599  {
600  std::cout << "Supported partition properties: ";
601  auto const part_props = device.get_info<sycl::info::device::partition_properties>();
602  auto has_affinity_domains = false;
603  for(auto const& prop : part_props)
604  {
605  switch(prop)
606  {
607  case sycl::info::partition_property::no_partition:
608  std::cout << "no partition";
609  break;
610 
611  case sycl::info::partition_property::partition_equally:
612  std::cout << "equally";
613  break;
614 
615  case sycl::info::partition_property::partition_by_counts:
616  std::cout << "by counts";
617  break;
618 
619  case sycl::info::partition_property::partition_by_affinity_domain:
620  std::cout << "by affinity domain";
621  has_affinity_domains = true;
622  break;
623 # if defined(BOOST_COMP_ICPX)
624  case sycl::info::partition_property::ext_intel_partition_by_cslice:
625  std::cout << "by compute slice (Intel extension; deprecated)";
626  break;
627 # endif
628  }
629  std::cout << ", ";
630  }
631  std::cout << '\n';
632 
633  if(has_affinity_domains)
634  {
635  std::cout << "Supported partition affinity domains: ";
636  auto const aff_doms = device.get_info<sycl::info::device::partition_affinity_domains>();
637  for(auto const& dom : aff_doms)
638  {
639  switch(dom)
640  {
641  case sycl::info::partition_affinity_domain::not_applicable:
642  std::cout << "not applicable";
643  break;
644 
645  case sycl::info::partition_affinity_domain::numa:
646  std::cout << "NUMA";
647  break;
648 
649  case sycl::info::partition_affinity_domain::L4_cache:
650  std::cout << "L4 cache";
651  break;
652 
653  case sycl::info::partition_affinity_domain::L3_cache:
654  std::cout << "L3 cache";
655  break;
656 
657  case sycl::info::partition_affinity_domain::L2_cache:
658  std::cout << "L2 cache";
659  break;
660 
661  case sycl::info::partition_affinity_domain::L1_cache:
662  std::cout << "L1 cache";
663  break;
664 
665  case sycl::info::partition_affinity_domain::next_partitionable:
666  std::cout << "next partitionable";
667  break;
668  }
669  std::cout << ", ";
670  }
671  std::cout << '\n';
672  }
673 
674  std::cout << "Current partition property: ";
675  switch(device.get_info<sycl::info::device::partition_type_property>())
676  {
677  case sycl::info::partition_property::no_partition:
678  std::cout << "no partition";
679  break;
680 
681  case sycl::info::partition_property::partition_equally:
682  std::cout << "partitioned equally";
683  break;
684 
685  case sycl::info::partition_property::partition_by_counts:
686  std::cout << "partitioned by counts";
687  break;
688 
689  case sycl::info::partition_property::partition_by_affinity_domain:
690  std::cout << "partitioned by affinity domain";
691  break;
692 
693 # if defined(BOOST_COMP_ICPX)
694  case sycl::info::partition_property::ext_intel_partition_by_cslice:
695  std::cout << "partitioned by compute slice (Intel extension; deprecated)";
696  break;
697 # endif
698  }
699  std::cout << '\n';
700 
701  std::cout << "Current partition affinity domain: ";
702  switch(device.get_info<sycl::info::device::partition_type_affinity_domain>())
703  {
704  case sycl::info::partition_affinity_domain::not_applicable:
705  std::cout << "not applicable";
706  break;
707 
708  case sycl::info::partition_affinity_domain::numa:
709  std::cout << "NUMA";
710  break;
711 
712  case sycl::info::partition_affinity_domain::L4_cache:
713  std::cout << "L4 cache";
714  break;
715 
716  case sycl::info::partition_affinity_domain::L3_cache:
717  std::cout << "L3 cache";
718  break;
719 
720  case sycl::info::partition_affinity_domain::L2_cache:
721  std::cout << "L2 cache";
722  break;
723 
724  case sycl::info::partition_affinity_domain::L1_cache:
725  std::cout << "L1 cache";
726  break;
727 
728  case sycl::info::partition_affinity_domain::next_partitionable:
729  std::cout << "next partitionable";
730  break;
731  }
732  std::cout << '\n';
733  }
734 
735  std::cout.flush();
736  }
737 # endif
738  };
739  } // namespace trait
740 } // namespace alpaka
741 
742 # if BOOST_COMP_CLANG
743 # pragma clang diagnostic pop
744 # endif
745 
746 #endif
ALPAKA_DEBUG_FULL_LOG_SCOPE
#define ALPAKA_DEBUG_FULL_LOG_SCOPE
Definition: Debug.hpp:62
alpaka::math::constants::e
constexpr double e
Definition: Traits.hpp:58
alpaka::math::fma
ALPAKA_NO_HOST_ACC_WARNING ALPAKA_FN_HOST_ACC auto fma(T const &fma_ctx, Tx const &x, Ty const &y, Tz const &z)
Computes x * y + z as if to infinite precision and rounded only once to fit the result type.
Definition: Traits.hpp:1134
alpaka::test::end
ALPAKA_FN_HOST auto end(TView &view) -> Iterator< TView >
Definition: Iterator.hpp:139
alpaka::test::begin
ALPAKA_FN_HOST auto begin(TView &view) -> Iterator< TView >
Definition: Iterator.hpp:133
alpaka::warp::all
ALPAKA_NO_HOST_ACC_WARNING ALPAKA_FN_ACC auto all(TWarp const &warp, std::int32_t predicate) -> std::int32_t
Evaluates predicate for all active threads of the warp and returns non-zero if and only if predicate ...
Definition: Traits.hpp:114
alpaka
The alpaka accelerator library.
Definition: AccCpuOmp2Blocks.hpp:49
alpaka::getDevCount
ALPAKA_FN_HOST auto getDevCount(TPlatform const &platform)
Definition: Traits.hpp:55
alpaka::getDevByIdx
ALPAKA_FN_HOST auto getDevByIdx(TPlatform const &platform, std::size_t const &devIdx) -> Dev< TPlatform >
Definition: Traits.hpp:62