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