Skip to content

[SYCL] Updated aspects list to be SYCL 2020 compliant #8751

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 10 commits into from
Mar 28, 2023
Merged

[SYCL] Updated aspects list to be SYCL 2020 compliant #8751

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
10 commits
Select commit Hold shift + click to select a range
b761381
Updated aspects list to be SYCL 2020 compliant
Feb 14, 2023
9a53480
Re-added symbols to prevent ABI break
Mar 23, 2023
a35e51e
Changes to host and usm_restricted_shared_allocations aspects
Mar 23, 2023
a0ebce0
Moved aspect::host deprecation test to correct location
Mar 24, 2023
48150ef
Merge branch 'sycl' into alamzeds/missing-device-aspects
Mar 24, 2023
1144e16
Removed redundant comments
Mar 24, 2023
7d62175
Removed deprecated aspects from tests
Mar 24, 2023
a69e298
Fixed deprecation message on host aspect
Mar 24, 2023
4800d46
Added TODO to implement emulated aspect for emulators
Mar 27, 2023
aba7fc1
Added missing newline to EOF
Mar 27, 2023
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
10 changes: 10 additions & 0 deletions sycl/include/sycl/device_aspect_macros.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -208,6 +208,11 @@
#define __SYCL_ALL_DEVICES_HAVE_39__ 0
#endif

#ifndef __SYCL_ALL_DEVICES_HAVE_40__
// __SYCL_ASPECT(emulated, 40)
#define __SYCL_ALL_DEVICES_HAVE_40__ 0
#endif

#ifndef __SYCL_ANY_DEVICE_HAS_0__
// __SYCL_ASPECT(host, 0)
#define __SYCL_ANY_DEVICE_HAS_0__ 0
Expand Down Expand Up @@ -407,3 +412,8 @@
// __SYCL_ASPECT(ext_intel_memory_bus_width, 39)
#define __SYCL_ANY_DEVICE_HAS_39__ 0
#endif

#ifndef __SYCL_ANY_DEVICE_HAS_40__
// __SYCL_ASPECT(emulated, 40)
#define __SYCL_ANY_DEVICE_HAS_40__ 0
#endif
6 changes: 6 additions & 0 deletions sycl/include/sycl/device_aspect_traits.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -132,6 +132,9 @@ struct all_devices_have<aspect::ext_intel_memory_clock_rate>
template <>
struct all_devices_have<aspect::ext_intel_memory_bus_width>
: std::bool_constant<__SYCL_ALL_DEVICES_HAVE_39__> {};
template <>
struct all_devices_have<aspect::emulated>
: std::bool_constant<__SYCL_ALL_DEVICES_HAVE_40__> {};

#ifdef __SYCL_ANY_DEVICE_HAS_ANY_ASPECT__
// Special case where any_device_has is trivially true.
Expand Down Expand Up @@ -258,6 +261,9 @@ struct any_device_has<aspect::ext_intel_memory_clock_rate>
template <>
struct any_device_has<aspect::ext_intel_memory_bus_width>
: std::bool_constant<__SYCL_ANY_DEVICE_HAS_39__> {};
template <>
struct any_device_has<aspect::emulated>
: std::bool_constant<__SYCL_ANY_DEVICE_HAS_40__> {};
#endif // __SYCL_ANY_DEVICE_HAS_ANY_ASPECT__

template <aspect Aspect>
Expand Down
3 changes: 1 addition & 2 deletions sycl/include/sycl/info/aspects.def
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -1,4 +1,3 @@
__SYCL_ASPECT(host, 0)
__SYCL_ASPECT(cpu, 1)
__SYCL_ASPECT(gpu, 2)
__SYCL_ASPECT(accelerator, 3)
Expand All @@ -12,7 +11,6 @@ __SYCL_ASPECT(queue_profiling, 12)
__SYCL_ASPECT(usm_device_allocations, 13)
__SYCL_ASPECT(usm_host_allocations, 14)
__SYCL_ASPECT(usm_shared_allocations, 15)
__SYCL_ASPECT(usm_restricted_shared_allocations, 16)
__SYCL_ASPECT(usm_system_allocations, 17)
__SYCL_ASPECT(ext_intel_pci_address, 18)
__SYCL_ASPECT(ext_intel_gpu_eu_count, 19)
Expand All @@ -36,3 +34,4 @@ __SYCL_ASPECT(ext_intel_free_memory, 36)
__SYCL_ASPECT(ext_intel_device_id, 37)
__SYCL_ASPECT(ext_intel_memory_clock_rate, 38)
__SYCL_ASPECT(ext_intel_memory_bus_width, 39)
__SYCL_ASPECT(emulated, 40)
2 changes: 2 additions & 0 deletions sycl/include/sycl/info/aspects_deprecated.def
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -2,3 +2,5 @@ __SYCL_ASPECT_DEPRECATED(int64_base_atomics, 7, "use atomic64 instead")
__SYCL_ASPECT_DEPRECATED(int64_extended_atomics, 8, "use atomic64 instead")
// Special macro for aspects that don't have own token
__SYCL_ASPECT_DEPRECATED_ALIAS(usm_system_allocator, usm_system_allocations, "use usm_system_allocations instead")
__SYCL_ASPECT_DEPRECATED(usm_restricted_shared_allocations, 16, "deprecated in SYCL 2020")
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

@gmlueck - It looks like this is in neither SYCL 1.2.1 nor SYCL 2020. Reading the definition of it, it does not seem to have a direct alternative. Is it okay to deprecate it (and remove it in an later release) or do we need an extension to have an alternative?

Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

This was an aspect we had in the SYCL 2020 provisional specification, but it was dropped by the time we finalized the spec. Yes, I think it's OK to deprecate it with no alternative. I think the original definition was never very clear.

__SYCL_ASPECT_DEPRECATED(host, 0, "removed in SYCL 2020, 'host' device has been removed")
5 changes: 5 additions & 0 deletions sycl/source/detail/device_impl.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -296,6 +296,11 @@ bool device_impl::has(aspect Aspect) const {
return is_accelerator();
case aspect::custom:
return false;
// TODO: Implement this for FPGA and ESIMD emulators.
case aspect::emulated:
return false;
case aspect::host_debuggable:
return false;
case aspect::fp16:
return has_extension("cl_khr_fp16");
case aspect::fp64:
Expand Down
93 changes: 43 additions & 50 deletions sycl/test/extensions/properties/properties_kernel.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -8,14 +8,12 @@ using namespace sycl::ext::oneapi::experimental;

using device_has_all =
decltype(device_has<
aspect::host, aspect::cpu, aspect::gpu, aspect::accelerator,
aspect::custom, aspect::fp16, aspect::fp64, aspect::image,
aspect::online_compiler, aspect::online_linker,
aspect::queue_profiling, aspect::usm_device_allocations,
aspect::usm_host_allocations, aspect::usm_shared_allocations,
aspect::usm_restricted_shared_allocations,
aspect::usm_system_allocations, aspect::ext_intel_pci_address,
aspect::ext_intel_gpu_eu_count,
aspect::cpu, aspect::gpu, aspect::accelerator, aspect::custom,
aspect::fp16, aspect::fp64, aspect::image, aspect::online_compiler,
aspect::online_linker, aspect::queue_profiling,
aspect::usm_device_allocations, aspect::usm_host_allocations,
aspect::usm_shared_allocations, aspect::usm_system_allocations,
aspect::ext_intel_pci_address, aspect::ext_intel_gpu_eu_count,
aspect::ext_intel_gpu_eu_simd_width, aspect::ext_intel_gpu_slices,
aspect::ext_intel_gpu_subslices_per_slice,
aspect::ext_intel_gpu_eu_count_per_subslice,
Expand Down Expand Up @@ -87,7 +85,6 @@ int main() {
static_assert(std::is_same_v<decltype(sub_group_size<28>)::value_t,
std::integral_constant<uint32_t, 28>>);

singleAspectDeviceHasChecks<aspect::host>();
singleAspectDeviceHasChecks<aspect::cpu>();
singleAspectDeviceHasChecks<aspect::gpu>();
singleAspectDeviceHasChecks<aspect::accelerator>();
Expand All @@ -101,7 +98,6 @@ int main() {
singleAspectDeviceHasChecks<aspect::usm_device_allocations>();
singleAspectDeviceHasChecks<aspect::usm_host_allocations>();
singleAspectDeviceHasChecks<aspect::usm_shared_allocations>();
singleAspectDeviceHasChecks<aspect::usm_restricted_shared_allocations>();
singleAspectDeviceHasChecks<aspect::usm_system_allocations>();
singleAspectDeviceHasChecks<aspect::ext_intel_pci_address>();
singleAspectDeviceHasChecks<aspect::ext_intel_gpu_eu_count>();
Expand Down Expand Up @@ -132,57 +128,54 @@ int main() {

static_assert(is_property_value<device_has_all>::value);
static_assert(std::is_same_v<device_has_key, device_has_all::key_t>);
static_assert(device_has_all::value.size() == 38);
static_assert(device_has_all::value[0] == aspect::host);
static_assert(device_has_all::value[1] == aspect::cpu);
static_assert(device_has_all::value[2] == aspect::gpu);
static_assert(device_has_all::value[3] == aspect::accelerator);
static_assert(device_has_all::value[4] == aspect::custom);
static_assert(device_has_all::value[5] == aspect::fp16);
static_assert(device_has_all::value[6] == aspect::fp64);
static_assert(device_has_all::value[7] == aspect::image);
static_assert(device_has_all::value[8] == aspect::online_compiler);
static_assert(device_has_all::value[9] == aspect::online_linker);
static_assert(device_has_all::value[10] == aspect::queue_profiling);
static_assert(device_has_all::value[11] == aspect::usm_device_allocations);
static_assert(device_has_all::value[12] == aspect::usm_host_allocations);
static_assert(device_has_all::value[13] == aspect::usm_shared_allocations);
static_assert(device_has_all::value[14] ==
aspect::usm_restricted_shared_allocations);
static_assert(device_has_all::value[15] == aspect::usm_system_allocations);
static_assert(device_has_all::value[16] == aspect::ext_intel_pci_address);
static_assert(device_has_all::value[17] == aspect::ext_intel_gpu_eu_count);
static_assert(device_has_all::value[18] ==
static_assert(device_has_all::value.size() == 36);
static_assert(device_has_all::value[0] == aspect::cpu);
static_assert(device_has_all::value[1] == aspect::gpu);
static_assert(device_has_all::value[2] == aspect::accelerator);
static_assert(device_has_all::value[3] == aspect::custom);
static_assert(device_has_all::value[4] == aspect::fp16);
static_assert(device_has_all::value[5] == aspect::fp64);
static_assert(device_has_all::value[6] == aspect::image);
static_assert(device_has_all::value[7] == aspect::online_compiler);
static_assert(device_has_all::value[8] == aspect::online_linker);
static_assert(device_has_all::value[9] == aspect::queue_profiling);
static_assert(device_has_all::value[10] == aspect::usm_device_allocations);
static_assert(device_has_all::value[11] == aspect::usm_host_allocations);
static_assert(device_has_all::value[12] == aspect::usm_shared_allocations);
static_assert(device_has_all::value[13] == aspect::usm_system_allocations);
static_assert(device_has_all::value[14] == aspect::ext_intel_pci_address);
static_assert(device_has_all::value[15] == aspect::ext_intel_gpu_eu_count);
static_assert(device_has_all::value[16] ==
aspect::ext_intel_gpu_eu_simd_width);
static_assert(device_has_all::value[19] == aspect::ext_intel_gpu_slices);
static_assert(device_has_all::value[20] ==
static_assert(device_has_all::value[17] == aspect::ext_intel_gpu_slices);
static_assert(device_has_all::value[18] ==
aspect::ext_intel_gpu_subslices_per_slice);
static_assert(device_has_all::value[21] ==
static_assert(device_has_all::value[19] ==
aspect::ext_intel_gpu_eu_count_per_subslice);
static_assert(device_has_all::value[22] ==
static_assert(device_has_all::value[20] ==
aspect::ext_intel_max_mem_bandwidth);
static_assert(device_has_all::value[23] == aspect::ext_intel_mem_channel);
static_assert(device_has_all::value[24] ==
static_assert(device_has_all::value[21] == aspect::ext_intel_mem_channel);
static_assert(device_has_all::value[22] ==
aspect::usm_atomic_host_allocations);
static_assert(device_has_all::value[25] ==
static_assert(device_has_all::value[23] ==
aspect::usm_atomic_shared_allocations);
static_assert(device_has_all::value[26] == aspect::atomic64);
static_assert(device_has_all::value[27] ==
static_assert(device_has_all::value[24] == aspect::atomic64);
static_assert(device_has_all::value[25] ==
aspect::ext_intel_device_info_uuid);
static_assert(device_has_all::value[28] == aspect::ext_oneapi_srgb);
static_assert(device_has_all::value[29] == aspect::ext_oneapi_native_assert);
static_assert(device_has_all::value[30] == aspect::host_debuggable);
static_assert(device_has_all::value[31] ==
static_assert(device_has_all::value[26] == aspect::ext_oneapi_srgb);
static_assert(device_has_all::value[27] == aspect::ext_oneapi_native_assert);
static_assert(device_has_all::value[28] == aspect::host_debuggable);
static_assert(device_has_all::value[29] ==
aspect::ext_intel_gpu_hw_threads_per_eu);
static_assert(device_has_all::value[32] ==
static_assert(device_has_all::value[30] ==
aspect::ext_oneapi_cuda_async_barrier);
static_assert(device_has_all::value[33] ==
static_assert(device_has_all::value[31] ==
aspect::ext_oneapi_bfloat16_math_functions);
static_assert(device_has_all::value[34] == aspect::ext_intel_free_memory);
static_assert(device_has_all::value[35] == aspect::ext_intel_device_id);
static_assert(device_has_all::value[36] ==
static_assert(device_has_all::value[32] == aspect::ext_intel_free_memory);
static_assert(device_has_all::value[33] == aspect::ext_intel_device_id);
static_assert(device_has_all::value[34] ==
aspect::ext_intel_memory_clock_rate);
static_assert(device_has_all::value[37] ==
static_assert(device_has_all::value[35] ==
aspect::ext_intel_memory_bus_width);
return 0;
}
13 changes: 5 additions & 8 deletions sycl/test/extensions/properties/properties_kernel_device_has.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -12,9 +12,8 @@ static constexpr auto device_has_all = device_has<
aspect::ext_oneapi_bfloat16_math_functions, aspect::custom, aspect::fp16,
aspect::fp64, aspect::image, aspect::online_compiler, aspect::online_linker,
aspect::queue_profiling, aspect::usm_device_allocations,
aspect::usm_restricted_shared_allocations, aspect::usm_system_allocations,
aspect::ext_intel_pci_address, aspect::host, aspect::cpu, aspect::gpu,
aspect::accelerator, aspect::ext_intel_gpu_eu_count,
aspect::usm_system_allocations, aspect::ext_intel_pci_address, aspect::cpu,
aspect::gpu, aspect::accelerator, aspect::ext_intel_gpu_eu_count,
aspect::ext_intel_gpu_subslices_per_slice,
aspect::ext_intel_gpu_eu_count_per_subslice,
aspect::ext_intel_max_mem_bandwidth, aspect::ext_intel_mem_channel,
Expand Down Expand Up @@ -141,10 +140,8 @@ int main() {
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"online_linker", i32 [[online_linker_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"queue_profiling", i32 [[queue_profiling_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"usm_device_allocations", i32 [[usm_device_allocations_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"usm_restricted_shared_allocations", i32 [[usm_restricted_shared_allocations_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"usm_system_allocations", i32 [[usm_system_allocations_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"ext_intel_pci_address", i32 [[ext_intel_pci_address_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"host", i32 [[host_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"cpu", i32 [[cpu_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"gpu", i32 [[gpu_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"accelerator", i32 [[accelerator_ASPECT_MD:[0-9]+]]}
Expand All @@ -168,6 +165,6 @@ int main() {
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"ext_intel_free_memory", i32 [[ext_intel_free_memory_ASPECT_MD:[0-9]+]]}
// CHECK-IR-DAG: !{{[0-9]+}} = !{!"ext_intel_device_id", i32 [[ext_intel_device_id_ASPECT_MD:[0-9]+]]}

// CHECK-IR-DAG: attributes #[[DHAttr1]] = { {{.*}}"sycl-device-has"="[[ext_oneapi_cuda_async_barrier_ASPECT_MD]],[[ext_oneapi_bfloat16_math_functions_ASPECT_MD]],[[custom_ASPECT_MD]],[[fp16_ASPECT_MD]],[[fp64_ASPECT_MD]],[[image_ASPECT_MD]],[[online_compiler_ASPECT_MD]],[[online_linker_ASPECT_MD]],[[queue_profiling_ASPECT_MD]],[[usm_device_allocations_ASPECT_MD]],[[usm_restricted_shared_allocations_ASPECT_MD]],[[usm_system_allocations_ASPECT_MD]],[[ext_intel_pci_address_ASPECT_MD]],[[host_ASPECT_MD]],[[cpu_ASPECT_MD]],[[gpu_ASPECT_MD]],[[accelerator_ASPECT_MD]],[[ext_intel_gpu_eu_count_ASPECT_MD]],[[ext_intel_gpu_subslices_per_slice_ASPECT_MD]],[[ext_intel_gpu_eu_count_per_subslice_ASPECT_MD]],[[ext_intel_max_mem_bandwidth_ASPECT_MD]],[[ext_intel_mem_channel_ASPECT_MD]],[[usm_atomic_host_allocations_ASPECT_MD]],[[usm_atomic_shared_allocations_ASPECT_MD]],[[atomic64_ASPECT_MD]],[[ext_intel_device_info_uuid_ASPECT_MD]],[[ext_oneapi_srgb_ASPECT_MD]],[[ext_intel_gpu_eu_simd_width_ASPECT_MD]],[[ext_intel_gpu_slices_ASPECT_MD]],[[ext_oneapi_native_assert_ASPECT_MD]],[[host_debuggable_ASPECT_MD]],[[ext_intel_gpu_hw_threads_per_eu_ASPECT_MD]],[[usm_host_allocations_ASPECT_MD]],[[usm_shared_allocations_ASPECT_MD]],[[ext_intel_free_memory_ASPECT_MD]],[[ext_intel_device_id_ASPECT_MD]]"
// CHECK-IR-DAG: attributes #[[DHAttr2]] = { {{.*}}"sycl-device-has"="[[ext_oneapi_cuda_async_barrier_ASPECT_MD]],[[ext_oneapi_bfloat16_math_functions_ASPECT_MD]],[[custom_ASPECT_MD]],[[fp16_ASPECT_MD]],[[fp64_ASPECT_MD]],[[image_ASPECT_MD]],[[online_compiler_ASPECT_MD]],[[online_linker_ASPECT_MD]],[[queue_profiling_ASPECT_MD]],[[usm_device_allocations_ASPECT_MD]],[[usm_restricted_shared_allocations_ASPECT_MD]],[[usm_system_allocations_ASPECT_MD]],[[ext_intel_pci_address_ASPECT_MD]],[[host_ASPECT_MD]],[[cpu_ASPECT_MD]],[[gpu_ASPECT_MD]],[[accelerator_ASPECT_MD]],[[ext_intel_gpu_eu_count_ASPECT_MD]],[[ext_intel_gpu_subslices_per_slice_ASPECT_MD]],[[ext_intel_gpu_eu_count_per_subslice_ASPECT_MD]],[[ext_intel_max_mem_bandwidth_ASPECT_MD]],[[ext_intel_mem_channel_ASPECT_MD]],[[usm_atomic_host_allocations_ASPECT_MD]],[[usm_atomic_shared_allocations_ASPECT_MD]],[[atomic64_ASPECT_MD]],[[ext_intel_device_info_uuid_ASPECT_MD]],[[ext_oneapi_srgb_ASPECT_MD]],[[ext_intel_gpu_eu_simd_width_ASPECT_MD]],[[ext_intel_gpu_slices_ASPECT_MD]],[[ext_oneapi_native_assert_ASPECT_MD]],[[host_debuggable_ASPECT_MD]],[[ext_intel_gpu_hw_threads_per_eu_ASPECT_MD]],[[usm_host_allocations_ASPECT_MD]],[[usm_shared_allocations_ASPECT_MD]],[[ext_intel_free_memory_ASPECT_MD]],[[ext_intel_device_id_ASPECT_MD]]"
// CHECK-IR-DAG: attributes #[[DHAttr3]] = { {{.*}}"sycl-device-has"="[[ext_oneapi_cuda_async_barrier_ASPECT_MD]],[[ext_oneapi_bfloat16_math_functions_ASPECT_MD]],[[custom_ASPECT_MD]],[[fp16_ASPECT_MD]],[[fp64_ASPECT_MD]],[[image_ASPECT_MD]],[[online_compiler_ASPECT_MD]],[[online_linker_ASPECT_MD]],[[queue_profiling_ASPECT_MD]],[[usm_device_allocations_ASPECT_MD]],[[usm_restricted_shared_allocations_ASPECT_MD]],[[usm_system_allocations_ASPECT_MD]],[[ext_intel_pci_address_ASPECT_MD]],[[host_ASPECT_MD]],[[cpu_ASPECT_MD]],[[gpu_ASPECT_MD]],[[accelerator_ASPECT_MD]],[[ext_intel_gpu_eu_count_ASPECT_MD]],[[ext_intel_gpu_subslices_per_slice_ASPECT_MD]],[[ext_intel_gpu_eu_count_per_subslice_ASPECT_MD]],[[ext_intel_max_mem_bandwidth_ASPECT_MD]],[[ext_intel_mem_channel_ASPECT_MD]],[[usm_atomic_host_allocations_ASPECT_MD]],[[usm_atomic_shared_allocations_ASPECT_MD]],[[atomic64_ASPECT_MD]],[[ext_intel_device_info_uuid_ASPECT_MD]],[[ext_oneapi_srgb_ASPECT_MD]],[[ext_intel_gpu_eu_simd_width_ASPECT_MD]],[[ext_intel_gpu_slices_ASPECT_MD]],[[ext_oneapi_native_assert_ASPECT_MD]],[[host_debuggable_ASPECT_MD]],[[ext_intel_gpu_hw_threads_per_eu_ASPECT_MD]],[[usm_host_allocations_ASPECT_MD]],[[usm_shared_allocations_ASPECT_MD]],[[ext_intel_free_memory_ASPECT_MD]],[[ext_intel_device_id_ASPECT_MD]]"
// CHECK-IR-DAG: attributes #[[DHAttr1]] = { {{.*}}"sycl-device-has"="[[ext_oneapi_cuda_async_barrier_ASPECT_MD]],[[ext_oneapi_bfloat16_math_functions_ASPECT_MD]],[[custom_ASPECT_MD]],[[fp16_ASPECT_MD]],[[fp64_ASPECT_MD]],[[image_ASPECT_MD]],[[online_compiler_ASPECT_MD]],[[online_linker_ASPECT_MD]],[[queue_profiling_ASPECT_MD]],[[usm_device_allocations_ASPECT_MD]],[[usm_system_allocations_ASPECT_MD]],[[ext_intel_pci_address_ASPECT_MD]],[[cpu_ASPECT_MD]],[[gpu_ASPECT_MD]],[[accelerator_ASPECT_MD]],[[ext_intel_gpu_eu_count_ASPECT_MD]],[[ext_intel_gpu_subslices_per_slice_ASPECT_MD]],[[ext_intel_gpu_eu_count_per_subslice_ASPECT_MD]],[[ext_intel_max_mem_bandwidth_ASPECT_MD]],[[ext_intel_mem_channel_ASPECT_MD]],[[usm_atomic_host_allocations_ASPECT_MD]],[[usm_atomic_shared_allocations_ASPECT_MD]],[[atomic64_ASPECT_MD]],[[ext_intel_device_info_uuid_ASPECT_MD]],[[ext_oneapi_srgb_ASPECT_MD]],[[ext_intel_gpu_eu_simd_width_ASPECT_MD]],[[ext_intel_gpu_slices_ASPECT_MD]],[[ext_oneapi_native_assert_ASPECT_MD]],[[host_debuggable_ASPECT_MD]],[[ext_intel_gpu_hw_threads_per_eu_ASPECT_MD]],[[usm_host_allocations_ASPECT_MD]],[[usm_shared_allocations_ASPECT_MD]],[[ext_intel_free_memory_ASPECT_MD]],[[ext_intel_device_id_ASPECT_MD]]"
// CHECK-IR-DAG: attributes #[[DHAttr2]] = { {{.*}}"sycl-device-has"="[[ext_oneapi_cuda_async_barrier_ASPECT_MD]],[[ext_oneapi_bfloat16_math_functions_ASPECT_MD]],[[custom_ASPECT_MD]],[[fp16_ASPECT_MD]],[[fp64_ASPECT_MD]],[[image_ASPECT_MD]],[[online_compiler_ASPECT_MD]],[[online_linker_ASPECT_MD]],[[queue_profiling_ASPECT_MD]],[[usm_device_allocations_ASPECT_MD]],[[usm_system_allocations_ASPECT_MD]],[[ext_intel_pci_address_ASPECT_MD]],[[cpu_ASPECT_MD]],[[gpu_ASPECT_MD]],[[accelerator_ASPECT_MD]],[[ext_intel_gpu_eu_count_ASPECT_MD]],[[ext_intel_gpu_subslices_per_slice_ASPECT_MD]],[[ext_intel_gpu_eu_count_per_subslice_ASPECT_MD]],[[ext_intel_max_mem_bandwidth_ASPECT_MD]],[[ext_intel_mem_channel_ASPECT_MD]],[[usm_atomic_host_allocations_ASPECT_MD]],[[usm_atomic_shared_allocations_ASPECT_MD]],[[atomic64_ASPECT_MD]],[[ext_intel_device_info_uuid_ASPECT_MD]],[[ext_oneapi_srgb_ASPECT_MD]],[[ext_intel_gpu_eu_simd_width_ASPECT_MD]],[[ext_intel_gpu_slices_ASPECT_MD]],[[ext_oneapi_native_assert_ASPECT_MD]],[[host_debuggable_ASPECT_MD]],[[ext_intel_gpu_hw_threads_per_eu_ASPECT_MD]],[[usm_host_allocations_ASPECT_MD]],[[usm_shared_allocations_ASPECT_MD]],[[ext_intel_free_memory_ASPECT_MD]],[[ext_intel_device_id_ASPECT_MD]]"
// CHECK-IR-DAG: attributes #[[DHAttr3]] = { {{.*}}"sycl-device-has"="[[ext_oneapi_cuda_async_barrier_ASPECT_MD]],[[ext_oneapi_bfloat16_math_functions_ASPECT_MD]],[[custom_ASPECT_MD]],[[fp16_ASPECT_MD]],[[fp64_ASPECT_MD]],[[image_ASPECT_MD]],[[online_compiler_ASPECT_MD]],[[online_linker_ASPECT_MD]],[[queue_profiling_ASPECT_MD]],[[usm_device_allocations_ASPECT_MD]],[[usm_system_allocations_ASPECT_MD]],[[ext_intel_pci_address_ASPECT_MD]],[[cpu_ASPECT_MD]],[[gpu_ASPECT_MD]],[[accelerator_ASPECT_MD]],[[ext_intel_gpu_eu_count_ASPECT_MD]],[[ext_intel_gpu_subslices_per_slice_ASPECT_MD]],[[ext_intel_gpu_eu_count_per_subslice_ASPECT_MD]],[[ext_intel_max_mem_bandwidth_ASPECT_MD]],[[ext_intel_mem_channel_ASPECT_MD]],[[usm_atomic_host_allocations_ASPECT_MD]],[[usm_atomic_shared_allocations_ASPECT_MD]],[[atomic64_ASPECT_MD]],[[ext_intel_device_info_uuid_ASPECT_MD]],[[ext_oneapi_srgb_ASPECT_MD]],[[ext_intel_gpu_eu_simd_width_ASPECT_MD]],[[ext_intel_gpu_slices_ASPECT_MD]],[[ext_oneapi_native_assert_ASPECT_MD]],[[host_debuggable_ASPECT_MD]],[[ext_intel_gpu_hw_threads_per_eu_ASPECT_MD]],[[usm_host_allocations_ASPECT_MD]],[[usm_shared_allocations_ASPECT_MD]],[[ext_intel_free_memory_ASPECT_MD]],[[ext_intel_device_id_ASPECT_MD]]"
Loading