[NFCI][SYCL] Refactor device_impl::get_info (#18358)

Before this change `get_info` was implemented via several layers:

1. `device::get_info` (ABI boundary)
2. `device_impl::get_info` (Some ABI exported for no reason!)
3. `detail::get_device_info`
4. `struct get_device_info_impl` 4b. specialization of 4. for SYCL/UR
types could be distinct, e.g.
    `get_device_info_impl<device, Desc>` calling
    `get_device_info_impl<ur_device_handle_t, Desc>`

with different descriptors implemented by partially/fully specializing
*either* 3 or 4 (with ABINeutral processing done at 1 as well).o

This PR collapses 2-4* into the following flow:

1. `device::get_info` (ABI boundary)
2. `device_impl::get_info`, single impl via `if constexpr` dispatch
3. (optional) for the most complicated descriptors outlined
implementation into dedicated named helpers with no template magic
4. Two uniform single-layer helpers `get_info_impl` and
`get_info_impl_nocheck`. A tiny bit of magic to support different types
(native UR/std::string/std::vector) is handled via `if constexpr` and is
easy to follow.

Old ABI exports are preserved with minor customizations based on preview
mode.

Different info descriptors are handled in the same order as they appear
in the corresponding `*.def` files.

I expect that having this new implementation in `device_impl.hpp` allows
for much more inlining without LTO, which previously wasn't possible
(some symbols resided in `device.cpp` while actual implementations were
only instantiated in `device_impl.cpp`).

Possible future changes based on top of this:

* Pre-initialize some bits in the `device_impl` ctor, at least
extensions string
* Possibly remove `UR_` mappings from `*.def` distributed as part of
headers (and maybe drop those files completely). This is an
implementation detail and has no business being visible to customers.
* Wider usage of `get_info_impl` helpers. At least some aspects code can
benefit from a follow-up cleanup.
* Use something like llvm::SmallString at the `device_impl` level

Author: aelovikov-intel

sycl/include/sycl/info/device_traits.def

@@ -135,6 +135,7 @@ __SYCL_PARAM_TRAITS_SPEC(device, execution_capabilities,
                          UR_DEVICE_INFO_EXECUTION_CAPABILITIES)
 __SYCL_PARAM_TRAITS_SPEC(device, queue_profiling, bool,
                          UR_DEVICE_INFO_QUEUE_PROPERTIES)
+// TODO: UR_DEVICE_INFO_FORCE_UINT32 looks wrong here:
 __SYCL_PARAM_TRAITS_SPEC(device, built_in_kernel_ids,
                          std::vector<sycl::kernel_id>, UR_DEVICE_INFO_FORCE_UINT32)
 __SYCL_PARAM_TRAITS_SPEC(device, built_in_kernels, std::vector<std::string>,

sycl/source/detail/allowlist.cpp

@@ -8,11 +8,11 @@
 #include <detail/allowlist.hpp>
 #include <detail/config.hpp>
 #include <detail/device_impl.hpp>
-#include <detail/device_info.hpp>
 #include <sycl/backend_types.hpp>
 
 #include <algorithm>
 #include <regex>
+#include <sstream>
 
 namespace sycl {
 inline namespace _V1 {
@@ -426,18 +426,18 @@ void applyAllowList(std::vector<ur_device_handle_t> &UrDevices,
     }
     // get DeviceVendorId value and put it to DeviceDesc
     uint32_t DeviceVendorIdUInt =
-        sycl::detail::get_device_info<info::device::vendor_id>(DeviceImpl);
+        DeviceImpl.get_info<info::device::vendor_id>();
     std::stringstream DeviceVendorIdHexStringStream;
     DeviceVendorIdHexStringStream << "0x" << std::hex << DeviceVendorIdUInt;
     const auto &DeviceVendorIdValue = DeviceVendorIdHexStringStream.str();
     DeviceDesc[DeviceVendorIdKeyName] = DeviceVendorIdValue;
     // get DriverVersion value and put it to DeviceDesc
     const std::string &DriverVersionValue =
-        sycl::detail::get_device_info<info::device::driver_version>(DeviceImpl);
+        DeviceImpl.get_info<info::device::driver_version>();
     DeviceDesc[DriverVersionKeyName] = DriverVersionValue;
     // get DeviceName value and put it to DeviceDesc
     const std::string &DeviceNameValue =
-        sycl::detail::get_device_info<info::device::name>(DeviceImpl);
+        DeviceImpl.get_info<info::device::name>();
     DeviceDesc[DeviceNameKeyName] = DeviceNameValue;
 
     // check if we can allow device with such device description DeviceDesc

sycl/source/detail/device_impl.cpp

@@ -7,7 +7,7 @@
 //===----------------------------------------------------------------------===//
 
 #include <detail/device_impl.hpp>
-#include <detail/device_info.hpp>
+#include <detail/jit_compiler.hpp>
 #include <detail/platform_impl.hpp>
 #include <detail/ur_info_code.hpp>
 #include <sycl/detail/ur.hpp>
@@ -78,30 +78,6 @@ platform device_impl::get_platform() const {
   return createSyclObjFromImpl<platform>(MPlatform);
 }
 
-template <typename Param>
-typename Param::return_type device_impl::get_info() const {
-  return get_device_info<Param>(*this);
-}
-// Explicitly instantiate all device info traits
-#define __SYCL_PARAM_TRAITS_SPEC(DescType, Desc, ReturnT, PiCode)              \
-  template ReturnT device_impl::get_info<info::device::Desc>() const;
-
-#define __SYCL_PARAM_TRAITS_SPEC_SPECIALIZED(DescType, Desc, ReturnT, PiCode)  \
-  template ReturnT device_impl::get_info<info::device::Desc>() const;
-
-#include <sycl/info/device_traits.def>
-#undef __SYCL_PARAM_TRAITS_SPEC_SPECIALIZED
-#undef __SYCL_PARAM_TRAITS_SPEC
-
-#define __SYCL_PARAM_TRAITS_SPEC(Namespace, DescType, Desc, ReturnT, PiCode)   \
-  template __SYCL_EXPORT ReturnT                                               \
-  device_impl::get_info<Namespace::info::DescType::Desc>() const;
-
-#include <sycl/info/ext_codeplay_device_traits.def>
-#include <sycl/info/ext_intel_device_traits.def>
-#include <sycl/info/ext_oneapi_device_traits.def>
-#undef __SYCL_PARAM_TRAITS_SPEC
-
 #ifndef __INTEL_PREVIEW_BREAKING_CHANGES
 template <>
 typename info::platform::version::return_type
@@ -146,7 +122,8 @@ device_impl::get_backend_info<info::device::backend_version>() const {
 
 bool device_impl::has_extension(const std::string &ExtensionName) const {
   std::string AllExtensionNames =
-      get_device_info_string(UR_DEVICE_INFO_EXTENSIONS);
+      get_info_impl<std::string>(UR_DEVICE_INFO_EXTENSIONS);
+
   return (AllExtensionNames.find(ExtensionName) != std::string::npos);
 }
 
@@ -256,6 +233,32 @@ device_impl::create_sub_devices(const std::vector<size_t> &Counts) const {
   return create_sub_devices(&Properties, Counts.size());
 }
 
+static inline std::string
+affinityDomainToString(info::partition_affinity_domain AffinityDomain) {
+  switch (AffinityDomain) {
+#define __SYCL_AFFINITY_DOMAIN_STRING_CASE(DOMAIN)                             \
+  case DOMAIN:                                                                 \
+    return #DOMAIN;
+
+    __SYCL_AFFINITY_DOMAIN_STRING_CASE(
+        sycl::info::partition_affinity_domain::numa)
+    __SYCL_AFFINITY_DOMAIN_STRING_CASE(
+        sycl::info::partition_affinity_domain::L4_cache)
+    __SYCL_AFFINITY_DOMAIN_STRING_CASE(
+        sycl::info::partition_affinity_domain::L3_cache)
+    __SYCL_AFFINITY_DOMAIN_STRING_CASE(
+        sycl::info::partition_affinity_domain::L2_cache)
+    __SYCL_AFFINITY_DOMAIN_STRING_CASE(
+        sycl::info::partition_affinity_domain::L1_cache)
+    __SYCL_AFFINITY_DOMAIN_STRING_CASE(
+        sycl::info::partition_affinity_domain::next_partitionable)
+#undef __SYCL_AFFINITY_DOMAIN_STRING_CASE
+  default:
+    assert(false && "Missing case for affinity domain.");
+    return "unknown";
+  }
+}
+
 std::vector<device> device_impl::create_sub_devices(
     info::partition_affinity_domain AffinityDomain) const {
   if (!is_partition_supported(
@@ -371,17 +374,15 @@ bool device_impl::has(aspect Aspect) const {
   case aspect::ext_oneapi_cuda_cluster_group:
     return get_info<info::device::ext_oneapi_cuda_cluster_group>();
   case aspect::usm_atomic_host_allocations:
-    return (
-        get_device_info_impl<ur_device_usm_access_capability_flags_t,
-                             info::device::usm_host_allocations>::get(*this) &
-        UR_DEVICE_USM_ACCESS_CAPABILITY_FLAG_ATOMIC_CONCURRENT_ACCESS);
+    return (get_info_impl<ur_device_usm_access_capability_flags_t>(
+                UR_DEVICE_INFO_USM_HOST_SUPPORT) &
+            UR_DEVICE_USM_ACCESS_CAPABILITY_FLAG_ATOMIC_CONCURRENT_ACCESS);
   case aspect::usm_shared_allocations:
     return get_info<info::device::usm_shared_allocations>();
   case aspect::usm_atomic_shared_allocations:
-    return (
-        get_device_info_impl<ur_device_usm_access_capability_flags_t,
-                             info::device::usm_shared_allocations>::get(*this) &
-        UR_DEVICE_USM_ACCESS_CAPABILITY_FLAG_ATOMIC_CONCURRENT_ACCESS);
+    return (get_info_impl<ur_device_usm_access_capability_flags_t>(
+                UR_DEVICE_INFO_USM_SINGLE_SHARED_SUPPORT) &
+            UR_DEVICE_USM_ACCESS_CAPABILITY_FLAG_ATOMIC_CONCURRENT_ACCESS);
   case aspect::usm_restricted_shared_allocations:
     return get_info<info::device::usm_restricted_shared_allocations>();
   case aspect::usm_system_allocations:
@@ -702,7 +703,7 @@ bool device_impl::has(aspect Aspect) const {
     return components.size() >= 2;
   }
   case aspect::ext_oneapi_is_component: {
-    typename sycl_to_ur<device>::type Result;
+    ur_device_handle_t Result;
     bool CallSuccessful =
         getAdapter()->call_nocheck<UrApiKind::urDeviceGetInfo>(
             getHandleRef(),
@@ -952,9 +953,7 @@ bool device_impl::supportsForwardProgress(
     ext::oneapi::experimental::forward_progress_guarantee guarantee,
     ext::oneapi::experimental::execution_scope threadScope,
     ext::oneapi::experimental::execution_scope coordinationScope) const {
-  using ReturnT =
-      std::vector<ext::oneapi::experimental::forward_progress_guarantee>;
-  auto guarantees = getProgressGuaranteesUpTo<ReturnT>(
+  auto guarantees = getProgressGuaranteesUpTo(
       getProgressGuarantee(threadScope, coordinationScope));
   return std::find(guarantees.begin(), guarantees.end(), guarantee) !=
          guarantees.end();
@@ -997,6 +996,62 @@ device_impl::getImmediateProgressGuarantee(
   return forward_progress_guarantee::weakly_parallel;
 }
 
+#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
+#define EXPORT_GET_INFO(PARAM)                                                 \
+  template <>                                                                  \
+  __SYCL_EXPORT PARAM::return_type device_impl::get_info<PARAM>() const {      \
+    return get_info_abi_workaround<PARAM>();                                   \
+  }
+
+// clang-format off
+EXPORT_GET_INFO(ext::intel::info::device::device_id)
+EXPORT_GET_INFO(ext::intel::info::device::pci_address)
+EXPORT_GET_INFO(ext::intel::info::device::gpu_eu_count)
+EXPORT_GET_INFO(ext::intel::info::device::gpu_eu_simd_width)
+EXPORT_GET_INFO(ext::intel::info::device::gpu_slices)
+EXPORT_GET_INFO(ext::intel::info::device::gpu_subslices_per_slice)
+EXPORT_GET_INFO(ext::intel::info::device::gpu_eu_count_per_subslice)
+EXPORT_GET_INFO(ext::intel::info::device::gpu_hw_threads_per_eu)
+EXPORT_GET_INFO(ext::intel::info::device::max_mem_bandwidth)
+EXPORT_GET_INFO(ext::intel::info::device::uuid)
+EXPORT_GET_INFO(ext::intel::info::device::free_memory)
+EXPORT_GET_INFO(ext::intel::info::device::memory_clock_rate)
+EXPORT_GET_INFO(ext::intel::info::device::memory_bus_width)
+EXPORT_GET_INFO(ext::intel::info::device::max_compute_queue_indices)
+EXPORT_GET_INFO(ext::intel::esimd::info::device::has_2d_block_io_support)
+EXPORT_GET_INFO(ext::intel::info::device::current_clock_throttle_reasons)
+EXPORT_GET_INFO(ext::intel::info::device::fan_speed)
+EXPORT_GET_INFO(ext::intel::info::device::min_power_limit)
+EXPORT_GET_INFO(ext::intel::info::device::max_power_limit)
+
+EXPORT_GET_INFO(ext::codeplay::experimental::info::device::supports_fusion)
+EXPORT_GET_INFO(ext::codeplay::experimental::info::device::max_registers_per_work_group)
+
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::max_global_work_groups)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::max_work_groups<1>)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::max_work_groups<2>)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::max_work_groups<3>)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::work_group_progress_capabilities<ext::oneapi::experimental::execution_scope::root_group>)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::sub_group_progress_capabilities<ext::oneapi::experimental::execution_scope::root_group>)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::sub_group_progress_capabilities<ext::oneapi::experimental::execution_scope::work_group>)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::work_item_progress_capabilities<ext::oneapi::experimental::execution_scope::root_group>)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::work_item_progress_capabilities<ext::oneapi::experimental::execution_scope::work_group>)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::work_item_progress_capabilities<ext::oneapi::experimental::execution_scope::sub_group>)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::architecture)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::matrix_combinations)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::image_row_pitch_align)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::max_image_linear_row_pitch)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::max_image_linear_width)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::max_image_linear_height)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::mipmap_max_anisotropy)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::component_devices)
+EXPORT_GET_INFO(ext::oneapi::experimental::info::device::composite_device)
+EXPORT_GET_INFO(ext::oneapi::info::device::num_compute_units)
+// clang-format on
+
+#undef EXPORT_GET_INFO
+#endif
+
 } // namespace detail
 } // namespace _V1
 } // namespace sycl