device_impl.cpp

Go to the documentation of this file.

//==----------------- device_impl.cpp - SYCL device ------------------------==//
//
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
// See https://llvm.org/LICENSE.txt for license information.
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
//
//===----------------------------------------------------------------------===//
 
#include <detail/device_impl.hpp>
#include <detail/device_info.hpp>
#include <detail/platform_impl.hpp>
#include <sycl/detail/ur.hpp>
#include <sycl/device.hpp>
 
#include <algorithm>
 
namespace sycl {
inline namespace _V1 {
namespace detail {
 
device_impl::device_impl(ur_native_handle_t InteropDeviceHandle,
                         const PluginPtr &Plugin)
    : device_impl(InteropDeviceHandle, nullptr, nullptr, Plugin) {}
 
device_impl::device_impl(ur_device_handle_t Device, PlatformImplPtr Platform)
    : device_impl(0, Device, Platform, Platform->getPlugin()) {}
 
device_impl::device_impl(ur_device_handle_t Device, const PluginPtr &Plugin)
    : device_impl(0, Device, nullptr, Plugin) {}
 
device_impl::device_impl(ur_native_handle_t InteropDeviceHandle,
                         ur_device_handle_t Device, PlatformImplPtr Platform,
                         const PluginPtr &Plugin)
    : MDevice(Device), MDeviceHostBaseTime(std::make_pair(0, 0)) {
  bool InteroperabilityConstructor = false;
  if (Device == nullptr) {
    assert(InteropDeviceHandle);
    // Get UR device from the raw device handle.
    // NOTE: this is for OpenCL interop only (and should go away).
    // With SYCL-2020 BE generalization "make" functions are used instead.
    Plugin->call(urDeviceCreateWithNativeHandle, InteropDeviceHandle, nullptr,
                 nullptr, &MDevice);
    InteroperabilityConstructor = true;
  }
 
  // TODO catch an exception and put it to list of asynchronous exceptions
  Plugin->call(urDeviceGetInfo, MDevice, UR_DEVICE_INFO_TYPE,
               sizeof(ur_device_type_t), &MType, nullptr);
 
  // No need to set MRootDevice when MAlwaysRootDevice is true
  if ((Platform == nullptr) || !Platform->MAlwaysRootDevice) {
    // TODO catch an exception and put it to list of asynchronous exceptions
    Plugin->call(urDeviceGetInfo, MDevice, UR_DEVICE_INFO_PARENT_DEVICE,
                 sizeof(ur_device_handle_t), &MRootDevice, nullptr);
  }
 
  if (!InteroperabilityConstructor) {
    // TODO catch an exception and put it to list of asynchronous exceptions
    // Interoperability Constructor already calls DeviceRetain in
    // urDeviceCreateWithNativeHandle.
    Plugin->call(urDeviceRetain, MDevice);
  }
 
  // set MPlatform
  if (!Platform) {
    Platform = platform_impl::getPlatformFromUrDevice(MDevice, Plugin);
  }
  MPlatform = Platform;
 
  MIsAssertFailSupported =
      has_extension(UR_DEVICE_INFO_EXTENSION_DEVICELIB_ASSERT);
}
 
device_impl::~device_impl() {
  // TODO catch an exception and put it to list of asynchronous exceptions
  const PluginPtr &Plugin = getPlugin();
  ur_result_t Err = Plugin->call_nocheck(urDeviceRelease, MDevice);
  __SYCL_CHECK_OCL_CODE_NO_EXC(Err);
}
 
bool device_impl::is_affinity_supported(
    info::partition_affinity_domain AffinityDomain) const {
  auto SupportedDomains = get_info<info::device::partition_affinity_domains>();
  return std::find(SupportedDomains.begin(), SupportedDomains.end(),
                   AffinityDomain) != SupportedDomains.end();
}
 
cl_device_id device_impl::get() const {
  // TODO catch an exception and put it to list of asynchronous exceptions
  getPlugin()->call(urDeviceRetain, MDevice);
  return ur::cast<cl_device_id>(getNative());
}
 
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>(
      MPlatform->getOrMakeDeviceImpl(MDevice, MPlatform));
}
// 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
 
template <>
typename info::platform::version::return_type
device_impl::get_backend_info<info::platform::version>() const {
  if (getBackend() != backend::opencl) {
    throw sycl::exception(errc::backend_mismatch,
                          "the info::platform::version info descriptor can "
                          "only be queried with an OpenCL backend");
  }
  return get_platform().get_info<info::platform::version>();
}
 
template <>
typename info::device::version::return_type
device_impl::get_backend_info<info::device::version>() const {
  if (getBackend() != backend::opencl) {
    throw sycl::exception(errc::backend_mismatch,
                          "the info::device::version info descriptor can only "
                          "be queried with an OpenCL backend");
  }
  return get_info<info::device::version>();
}
 
template <>
typename info::device::backend_version::return_type
device_impl::get_backend_info<info::device::backend_version>() const {
  if (getBackend() != backend::ext_oneapi_level_zero) {
    throw sycl::exception(errc::backend_mismatch,
                          "the info::device::backend_version info descriptor "
                          "can only be queried with a Level Zero backend");
  }
  return "";
  // Currently The Level Zero backend does not define the value of this
  // information descriptor and implementations are encouraged to return the
  // empty string as per specification.
}
 
bool device_impl::has_extension(const std::string &ExtensionName) const {
  std::string AllExtensionNames =
      get_device_info_string(UR_DEVICE_INFO_EXTENSIONS);
  return (AllExtensionNames.find(ExtensionName) != std::string::npos);
}
 
bool device_impl::is_partition_supported(info::partition_property Prop) const {
  auto SupportedProperties = get_info<info::device::partition_properties>();
  return std::find(SupportedProperties.begin(), SupportedProperties.end(),
                   Prop) != SupportedProperties.end();
}
 
std::vector<device> device_impl::create_sub_devices(
    const ur_device_partition_properties_t *Properties,
    size_t SubDevicesCount) const {
  std::vector<ur_device_handle_t> SubDevices(SubDevicesCount);
  uint32_t ReturnedSubDevices = 0;
  const PluginPtr &Plugin = getPlugin();
  Plugin->call<sycl::errc::invalid>(urDevicePartition, MDevice, Properties,
                                    SubDevicesCount, SubDevices.data(),
                                    &ReturnedSubDevices);
  if (ReturnedSubDevices != SubDevicesCount) {
    throw sycl::exception(
        errc::invalid,
        "Could not partition to the specified number of sub-devices");
  }
  // TODO: Need to describe the subdevice model. Some sub_device management
  // may be necessary. What happens if create_sub_devices is called multiple
  // times with the same arguments?
  //
  std::vector<device> res;
  std::for_each(SubDevices.begin(), SubDevices.end(),
                [&res, this](const ur_device_handle_t &a_ur_device) {
                  device sycl_device = detail::createSyclObjFromImpl<device>(
                      MPlatform->getOrMakeDeviceImpl(a_ur_device, MPlatform));
                  res.push_back(sycl_device);
                });
  return res;
}
 
std::vector<device> device_impl::create_sub_devices(size_t ComputeUnits) const {
  if (!is_partition_supported(info::partition_property::partition_equally)) {
    throw sycl::exception(make_error_code(errc::feature_not_supported),
                          "Device does not support "
                          "sycl::info::partition_property::partition_equally.");
  }
  // If count exceeds the total number of compute units in the device, an
  // exception with the errc::invalid error code must be thrown.
  auto MaxComputeUnits = get_info<info::device::max_compute_units>();
  if (ComputeUnits > MaxComputeUnits)
    throw sycl::exception(errc::invalid,
                          "Total counts exceed max compute units");
 
  size_t SubDevicesCount = MaxComputeUnits / ComputeUnits;
 
  ur_device_partition_property_t Prop{};
  Prop.type = UR_DEVICE_PARTITION_EQUALLY;
  Prop.value.count = static_cast<uint32_t>(ComputeUnits);
 
  ur_device_partition_properties_t Properties{};
  Properties.stype = UR_STRUCTURE_TYPE_DEVICE_PARTITION_PROPERTIES;
  Properties.PropCount = 1;
  Properties.pProperties = &Prop;
 
  return create_sub_devices(&Properties, SubDevicesCount);
}
 
std::vector<device>
device_impl::create_sub_devices(const std::vector<size_t> &Counts) const {
  if (!is_partition_supported(info::partition_property::partition_by_counts)) {
    throw sycl::exception(
        make_error_code(errc::feature_not_supported),
        "Device does not support "
        "sycl::info::partition_property::partition_by_counts.");
  }
 
  std::vector<ur_device_partition_property_t> Props{};
 
  // Fill the properties vector with counts and validate it
  size_t TotalCounts = 0;
  size_t NonZeroCounts = 0;
  for (auto Count : Counts) {
    TotalCounts += Count;
    NonZeroCounts += (Count != 0) ? 1 : 0;
    Props.push_back(ur_device_partition_property_t{
        UR_DEVICE_PARTITION_BY_COUNTS, {static_cast<uint32_t>(Count)}});
  }
 
  ur_device_partition_properties_t Properties{};
  Properties.stype = UR_STRUCTURE_TYPE_DEVICE_PARTITION_PROPERTIES;
  Properties.pProperties = Props.data();
  Properties.PropCount = Props.size();
 
  // If the number of non-zero values in counts exceeds the device’s maximum
  // number of sub devices (as returned by info::device::
  // partition_max_sub_devices) an exception with the errc::invalid
  // error code must be thrown.
  if (NonZeroCounts > get_info<info::device::partition_max_sub_devices>())
    throw sycl::exception(errc::invalid,
                          "Total non-zero counts exceed max sub-devices");
 
  // If the total of all the values in the counts vector exceeds the total
  // number of compute units in the device (as returned by
  // info::device::max_compute_units), an exception with the errc::invalid
  // error code must be thrown.
  if (TotalCounts > get_info<info::device::max_compute_units>())
    throw sycl::exception(errc::invalid,
                          "Total counts exceed max compute units");
 
  return create_sub_devices(&Properties, Counts.size());
}
 
std::vector<device> device_impl::create_sub_devices(
    info::partition_affinity_domain AffinityDomain) const {
  if (!is_partition_supported(
          info::partition_property::partition_by_affinity_domain)) {
    throw sycl::exception(
        make_error_code(errc::feature_not_supported),
        "Device does not support "
        "sycl::info::partition_property::partition_by_affinity_domain.");
  }
  if (!is_affinity_supported(AffinityDomain)) {
    throw sycl::exception(make_error_code(errc::feature_not_supported),
                          "Device does not support " +
                              affinityDomainToString(AffinityDomain) + ".");
  }
 
  ur_device_partition_property_t Prop;
  Prop.type = UR_DEVICE_PARTITION_BY_AFFINITY_DOMAIN;
  Prop.value.affinity_domain =
      static_cast<ur_device_affinity_domain_flags_t>(AffinityDomain);
 
  ur_device_partition_properties_t Properties{};
  Properties.stype = UR_STRUCTURE_TYPE_DEVICE_PARTITION_PROPERTIES;
  Properties.PropCount = 1;
  Properties.pProperties = &Prop;
 
  uint32_t SubDevicesCount = 0;
  const PluginPtr &Plugin = getPlugin();
  Plugin->call<sycl::errc::invalid>(urDevicePartition, MDevice, &Properties, 0,
                                    nullptr, &SubDevicesCount);
 
  return create_sub_devices(&Properties, SubDevicesCount);
}
 
std::vector<device> device_impl::create_sub_devices() const {
  if (!is_partition_supported(
          info::partition_property::ext_intel_partition_by_cslice)) {
    throw sycl::exception(
        make_error_code(errc::feature_not_supported),
        "Device does not support "
        "sycl::info::partition_property::ext_intel_partition_by_cslice.");
  }
 
  ur_device_partition_property_t Prop;
  Prop.type = UR_DEVICE_PARTITION_BY_CSLICE;
 
  ur_device_partition_properties_t Properties{};
  Properties.stype = UR_STRUCTURE_TYPE_DEVICE_PARTITION_PROPERTIES;
  Properties.pProperties = &Prop;
  Properties.PropCount = 1;
 
  uint32_t SubDevicesCount = 0;
  const PluginPtr &Plugin = getPlugin();
  Plugin->call(urDevicePartition, MDevice, &Properties, 0, nullptr,
               &SubDevicesCount);
 
  return create_sub_devices(&Properties, SubDevicesCount);
}
 
ur_native_handle_t device_impl::getNative() const {
  auto Plugin = getPlugin();
  if (getBackend() == backend::opencl)
    Plugin->call(urDeviceRetain, getHandleRef());
  ur_native_handle_t Handle;
  Plugin->call(urDeviceGetNativeHandle, getHandleRef(), &Handle);
  return Handle;
}
 
bool device_impl::has(aspect Aspect) const {
  size_t return_size = 0;
 
  switch (Aspect) {
  case aspect::host:
    // Deprecated
    return false;
  case aspect::cpu:
    return is_cpu();
  case aspect::gpu:
    return is_gpu();
  case aspect::accelerator:
    return is_accelerator();
  case aspect::custom:
    return false;
  // TODO: Implement this for FPGA emulator.
  case aspect::emulated:
    return false;
  case aspect::host_debuggable:
    return false;
  case aspect::fp16:
    return has_extension("cl_khr_fp16");
  case aspect::fp64:
    return has_extension("cl_khr_fp64");
  case aspect::int64_base_atomics:
    return has_extension("cl_khr_int64_base_atomics");
  case aspect::int64_extended_atomics:
    return has_extension("cl_khr_int64_extended_atomics");
  case aspect::atomic64:
    return get_info<info::device::atomic64>();
  case aspect::image:
    return get_info<info::device::image_support>();
  case aspect::online_compiler:
    return get_info<info::device::is_compiler_available>();
  case aspect::online_linker:
    return get_info<info::device::is_linker_available>();
  case aspect::queue_profiling:
    return get_info<info::device::queue_profiling>();
  case aspect::usm_device_allocations:
    return get_info<info::device::usm_device_allocations>();
  case aspect::usm_host_allocations:
    return get_info<info::device::usm_host_allocations>();
  case aspect::ext_intel_mem_channel:
    return get_info<info::device::ext_intel_mem_channel>();
  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(MPlatform->getDeviceImpl(MDevice)) &
            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(MPlatform->getDeviceImpl(MDevice)) &
            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:
    return get_info<info::device::usm_system_allocations>();
  case aspect::ext_intel_device_id:
    return getPlugin()->call_nocheck(urDeviceGetInfo, MDevice,
                                     UR_DEVICE_INFO_DEVICE_ID, 0, nullptr,
                                     &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_pci_address:
    return getPlugin()->call_nocheck(urDeviceGetInfo, MDevice,
                                     UR_DEVICE_INFO_PCI_ADDRESS, 0, nullptr,
                                     &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_gpu_eu_count:
    return getPlugin()->call_nocheck(urDeviceGetInfo, MDevice,
                                     UR_DEVICE_INFO_GPU_EU_COUNT, 0, nullptr,
                                     &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_gpu_eu_simd_width:
    return getPlugin()->call_nocheck(
               urDeviceGetInfo, MDevice, UR_DEVICE_INFO_GPU_EU_SIMD_WIDTH, 0,
               nullptr, &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_gpu_slices:
    return getPlugin()->call_nocheck(urDeviceGetInfo, MDevice,
                                     UR_DEVICE_INFO_GPU_EU_SLICES, 0, nullptr,
                                     &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_gpu_subslices_per_slice:
    return getPlugin()->call_nocheck(
               urDeviceGetInfo, MDevice, UR_DEVICE_INFO_GPU_SUBSLICES_PER_SLICE,
               0, nullptr, &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_gpu_eu_count_per_subslice:
    return getPlugin()->call_nocheck(urDeviceGetInfo, MDevice,
                                     UR_DEVICE_INFO_GPU_EU_COUNT_PER_SUBSLICE,
                                     0, nullptr,
                                     &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_gpu_hw_threads_per_eu:
    return getPlugin()->call_nocheck(
               urDeviceGetInfo, MDevice, UR_DEVICE_INFO_GPU_HW_THREADS_PER_EU,
               0, nullptr, &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_free_memory:
    return getPlugin()->call_nocheck(urDeviceGetInfo, MDevice,
                                     UR_DEVICE_INFO_GLOBAL_MEM_FREE, 0, nullptr,
                                     &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_memory_clock_rate:
    return getPlugin()->call_nocheck(
               urDeviceGetInfo, MDevice, UR_DEVICE_INFO_MEMORY_CLOCK_RATE, 0,
               nullptr, &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_memory_bus_width:
    return getPlugin()->call_nocheck(
               urDeviceGetInfo, MDevice, UR_DEVICE_INFO_MEMORY_BUS_WIDTH, 0,
               nullptr, &return_size) == UR_RESULT_SUCCESS;
  case aspect::ext_intel_device_info_uuid: {
    auto Result =
        getPlugin()->call_nocheck(urDeviceGetInfo, MDevice, UR_DEVICE_INFO_UUID,
                                  0, nullptr, &return_size);
    if (Result != UR_RESULT_SUCCESS) {
      return false;
    }
 
    assert(return_size <= 16);
    unsigned char UUID[16];
 
    return getPlugin()->call_nocheck(
               urDeviceGetInfo, MDevice, UR_DEVICE_INFO_UUID,
               16 * sizeof(unsigned char), UUID, nullptr) == UR_RESULT_SUCCESS;
  }
  case aspect::ext_intel_max_mem_bandwidth:
    // currently not supported
    return false;
  case aspect::ext_oneapi_srgb:
    return get_info<info::device::ext_oneapi_srgb>();
  case aspect::ext_oneapi_native_assert:
    return isAssertFailSupported();
  case aspect::ext_oneapi_cuda_async_barrier: {
    int async_barrier_supported;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice, UR_DEVICE_INFO_ASYNC_BARRIER, sizeof(int),
            &async_barrier_supported, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && async_barrier_supported;
  }
  case aspect::ext_intel_legacy_image: {
    ur_bool_t legacy_image_support = false;
    bool call_successful =
        getPlugin()->call_nocheck(urDeviceGetInfo, MDevice,
                                  UR_DEVICE_INFO_IMAGE_SUPPORTED,
                                  sizeof(ur_bool_t), &legacy_image_support,
                                  nullptr) == UR_RESULT_SUCCESS;
    return call_successful && legacy_image_support;
  }
  case aspect::ext_oneapi_bindless_images: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(urDeviceGetInfo, MDevice,
                                  UR_DEVICE_INFO_BINDLESS_IMAGES_SUPPORT_EXP,
                                  sizeof(ur_bool_t), &support,
                                  nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_images_shared_usm: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_IMAGES_SHARED_USM_SUPPORT_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_images_1d_usm: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_IMAGES_1D_USM_SUPPORT_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_images_2d_usm: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_IMAGES_2D_USM_SUPPORT_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_interop_memory_import: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_INTEROP_MEMORY_IMPORT_SUPPORT_EXP, sizeof(ur_bool_t),
            &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_interop_semaphore_import: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_INTEROP_SEMAPHORE_IMPORT_SUPPORT_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_mipmap: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice, UR_DEVICE_INFO_MIPMAP_SUPPORT_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_mipmap_anisotropy: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(urDeviceGetInfo, MDevice,
                                  UR_DEVICE_INFO_MIPMAP_ANISOTROPY_SUPPORT_EXP,
                                  sizeof(ur_bool_t), &support,
                                  nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_mipmap_level_reference: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_MIPMAP_LEVEL_REFERENCE_SUPPORT_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_sampled_image_fetch_1d_usm: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_1D_USM_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_sampled_image_fetch_1d: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_1D_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_sampled_image_fetch_2d_usm: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_2D_USM_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_sampled_image_fetch_2d: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_2D_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_sampled_image_fetch_3d: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_3D_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_cubemap: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice, UR_DEVICE_INFO_CUBEMAP_SUPPORT_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_cubemap_seamless_filtering: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_CUBEMAP_SEAMLESS_FILTERING_SUPPORT_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_image_array: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice, UR_DEVICE_INFO_IMAGE_ARRAY_SUPPORT_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_unique_addressing_per_dim: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_UNIQUE_ADDRESSING_PER_DIM_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_images_sample_1d_usm: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_1D_USM_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_bindless_images_sample_2d_usm: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_2D_USM_EXP,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_intel_esimd: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice, UR_DEVICE_INFO_ESIMD_SUPPORT,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_ballot_group:
  case aspect::ext_oneapi_fixed_size_group:
  case aspect::ext_oneapi_opportunistic_group: {
    return (this->getBackend() == backend::ext_oneapi_level_zero) ||
           (this->getBackend() == backend::opencl) ||
           (this->getBackend() == backend::ext_oneapi_cuda);
  }
  case aspect::ext_oneapi_tangle_group: {
    // TODO: tangle_group is not currently supported for CUDA devices. Add when
    //       implemented.
    return (this->getBackend() == backend::ext_oneapi_level_zero) ||
           (this->getBackend() == backend::opencl);
  }
  case aspect::ext_intel_matrix: {
    using arch = sycl::ext::oneapi::experimental::architecture;
    const std::vector<arch> supported_archs = {
        arch::intel_cpu_spr,     arch::intel_cpu_gnr,
        arch::intel_gpu_pvc,     arch::intel_gpu_dg2_g10,
        arch::intel_gpu_dg2_g11, arch::intel_gpu_dg2_g12};
    try {
      return std::any_of(
          supported_archs.begin(), supported_archs.end(),
          [=](const arch a) { return this->extOneapiArchitectureIs(a); });
    } catch (const sycl::exception &) {
      // If we're here it means the device does not support architecture
      // querying
      return false;
    }
  }
  case aspect::ext_oneapi_is_composite: {
    auto components = get_info<
        sycl::ext::oneapi::experimental::info::device::component_devices>();
    // Any device with ext_oneapi_is_composite aspect will have at least two
    // constituent component devices.
    return components.size() >= 2;
  }
  case aspect::ext_oneapi_is_component: {
    typename sycl_to_ur<device>::type Result;
    bool CallSuccessful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, getHandleRef(),
            UrInfoCode<ext::oneapi::experimental::info::device::
                           composite_device>::value,
            sizeof(Result), &Result, nullptr) == UR_RESULT_SUCCESS;
 
    return CallSuccessful && Result != nullptr;
  }
  case aspect::ext_oneapi_graph: {
    bool SupportsCommandBufferUpdate = false;
    bool CallSuccessful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_COMMAND_BUFFER_UPDATE_SUPPORT_EXP,
            sizeof(SupportsCommandBufferUpdate), &SupportsCommandBufferUpdate,
            nullptr) == UR_RESULT_SUCCESS;
    if (!CallSuccessful) {
      return false;
    }
 
    return has(aspect::ext_oneapi_limited_graph) && SupportsCommandBufferUpdate;
  }
  case aspect::ext_oneapi_limited_graph: {
    bool SupportsCommandBuffers = false;
    bool CallSuccessful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice, UR_DEVICE_INFO_COMMAND_BUFFER_SUPPORT_EXP,
            sizeof(SupportsCommandBuffers), &SupportsCommandBuffers,
            nullptr) == UR_RESULT_SUCCESS;
    if (!CallSuccessful) {
      return false;
    }
 
    return SupportsCommandBuffers;
  }
  case aspect::ext_oneapi_private_alloca: {
    // Extension only supported on SPIR-V targets.
    backend be = getBackend();
    return be == sycl::backend::ext_oneapi_level_zero ||
           be == sycl::backend::opencl;
  }
  case aspect::ext_oneapi_queue_profiling_tag: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice,
            UR_DEVICE_INFO_TIMESTAMP_RECORDING_SUPPORT_EXP, sizeof(ur_bool_t),
            &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_oneapi_virtual_mem: {
    ur_bool_t support = false;
    bool call_successful =
        getPlugin()->call_nocheck(
            urDeviceGetInfo, MDevice, UR_DEVICE_INFO_VIRTUAL_MEMORY_SUPPORT,
            sizeof(ur_bool_t), &support, nullptr) == UR_RESULT_SUCCESS;
    return call_successful && support;
  }
  case aspect::ext_intel_fpga_task_sequence: {
    return is_accelerator();
  }
  }
 
  return false; // This device aspect has not been implemented yet.
}
 
bool device_impl::isAssertFailSupported() const {
  return MIsAssertFailSupported;
}
 
std::string device_impl::getDeviceName() const {
  std::call_once(MDeviceNameFlag,
                 [this]() { MDeviceName = get_info<info::device::name>(); });
 
  return MDeviceName;
}
 
ext::oneapi::experimental::architecture device_impl::getDeviceArch() const {
  std::call_once(MDeviceArchFlag, [this]() {
    MDeviceArch =
        get_info<ext::oneapi::experimental::info::device::architecture>();
  });
 
  return MDeviceArch;
}
 
// On the first call this function queries for device timestamp
// along with host synchronized timestamp and stores it in member variable
// MDeviceHostBaseTime. Subsequent calls to this function would just retrieve
// the host timestamp, compute difference against the host timestamp in
// MDeviceHostBaseTime and calculate the device timestamp based on the
// difference.
//
// The MDeviceHostBaseTime is refreshed with new device and host timestamp
// after a certain interval (determined by TimeTillRefresh) to account for
// clock drift between host and device.
//
uint64_t device_impl::getCurrentDeviceTime() {
  using namespace std::chrono;
  uint64_t HostTime =
      duration_cast<nanoseconds>(steady_clock::now().time_since_epoch())
          .count();
 
  // To account for potential clock drift between host clock and device clock.
  // The value set is arbitrary: 200 seconds
  constexpr uint64_t TimeTillRefresh = 200e9;
  assert(HostTime >= MDeviceHostBaseTime.second);
  uint64_t Diff = HostTime - MDeviceHostBaseTime.second;
 
  // If getCurrentDeviceTime is called for the first time or we have to refresh.
  if (!MDeviceHostBaseTime.second || Diff > TimeTillRefresh) {
    const auto &Plugin = getPlugin();
    auto Result = Plugin->call_nocheck(urDeviceGetGlobalTimestamps, MDevice,
                                       &MDeviceHostBaseTime.first,
                                       &MDeviceHostBaseTime.second);
    // We have to remember base host timestamp right after UR call and it is
    // going to be used for calculation of the device timestamp at the next
    // getCurrentDeviceTime() call. We need to do it here because getPlugin()
    // and urDeviceGetGlobalTimestamps calls may take significant amount of time,
    // for example on the first call to getPlugin plugins may need to be
    // initialized. If we use timestamp from the beginning of the function then
    // the difference between host timestamps of the current
    // getCurrentDeviceTime and the next getCurrentDeviceTime will be incorrect
    // because it will include execution time of the code before we get device
    // timestamp from urDeviceGetGlobalTimestamps.
    HostTime =
        duration_cast<nanoseconds>(steady_clock::now().time_since_epoch())
            .count();
    if (Result == UR_RESULT_ERROR_INVALID_OPERATION) {
      // NOTE(UR port): Removed the call to GetLastError because  we shouldn't
      // be calling it after ERROR_INVALID_OPERATION: there is no
      // adapter-specific error.
      throw detail::set_ur_error(
          sycl::exception(
              make_error_code(errc::feature_not_supported),
              "Device and/or backend does not support querying timestamp."),
          UR_RESULT_ERROR_INVALID_OPERATION);
    } else {
      Plugin->checkUrResult<errc::feature_not_supported>(Result);
    }
    // Until next sync we will compute device time based on the host time
    // returned in HostTime, so make this our base host time.
    MDeviceHostBaseTime.second = HostTime;
    Diff = 0;
  }
  return MDeviceHostBaseTime.first + Diff;
}
 
bool device_impl::isGetDeviceAndHostTimerSupported() {
  const auto &Plugin = getPlugin();
  uint64_t DeviceTime = 0, HostTime = 0;
  auto Result = Plugin->call_nocheck(urDeviceGetGlobalTimestamps, MDevice,
                                     &DeviceTime, &HostTime);
  return Result != UR_RESULT_ERROR_INVALID_OPERATION;
}
 
bool device_impl::extOneapiCanCompile(
    ext::oneapi::experimental::source_language Language) {
  try {
    return sycl::ext::oneapi::experimental::detail::
        is_source_kernel_bundle_supported(getBackend(), Language);
  } catch (sycl::exception &) {
    return false;
  }
}
 
// Returns the strongest guarantee that can be provided by the host device for
// threads created at threadScope from a coordination scope given by
// coordinationScope
sycl::ext::oneapi::experimental::forward_progress_guarantee
device_impl::getHostProgressGuarantee(
    ext::oneapi::experimental::execution_scope,
    ext::oneapi::experimental::execution_scope) {
  return sycl::ext::oneapi::experimental::forward_progress_guarantee::
      weakly_parallel;
}
 
// Returns the strongest progress guarantee that can be provided by this device
// for threads created at threadScope from the coordination scope given by
// coordinationScope.
sycl::ext::oneapi::experimental::forward_progress_guarantee
device_impl::getProgressGuarantee(
    ext::oneapi::experimental::execution_scope threadScope,
    ext::oneapi::experimental::execution_scope coordinationScope) const {
  using forward_progress_guarantee =
      ext::oneapi::experimental::forward_progress_guarantee;
  using execution_scope = ext::oneapi::experimental::execution_scope;
  const int executionScopeSize = 4;
  (void)coordinationScope;
  int threadScopeNum = static_cast<int>(threadScope);
  // we get the immediate progress guarantee that is provided by each scope
  // between root_group and threadScope and then return the weakest of these.
  // Counterintuitively, this corresponds to taking the max of the enum values
  // because of how the forward_progress_guarantee enum values are declared.
  int guaranteeNum = static_cast<int>(
      getImmediateProgressGuarantee(execution_scope::root_group));
  for (int currentScope = executionScopeSize - 2; currentScope > threadScopeNum;
       --currentScope) {
    guaranteeNum = std::max(guaranteeNum,
                            static_cast<int>(getImmediateProgressGuarantee(
                                static_cast<execution_scope>(currentScope))));
  }
  return static_cast<forward_progress_guarantee>(guaranteeNum);
}
 
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>(
      getProgressGuarantee(threadScope, coordinationScope));
  return std::find(guarantees.begin(), guarantees.end(), guarantee) !=
         guarantees.end();
}
 
// Returns the progress guarantee provided for a coordination scope
// given by coordination_scope for threads created at a scope
// immediately below coordination_scope. For example, for root_group
// coordination scope it returns the progress guarantee provided
// at root_group for threads created at work_group.
ext::oneapi::experimental::forward_progress_guarantee
device_impl::getImmediateProgressGuarantee(
    ext::oneapi::experimental::execution_scope coordination_scope) const {
  using forward_progress_guarantee =
      ext::oneapi::experimental::forward_progress_guarantee;
  using execution_scope = ext::oneapi::experimental::execution_scope;
  if (is_cpu() && getBackend() == backend::opencl) {
    switch (coordination_scope) {
    case execution_scope::root_group:
      return forward_progress_guarantee::parallel;
    case execution_scope::work_group:
    case execution_scope::sub_group:
      return forward_progress_guarantee::weakly_parallel;
    default:
      throw sycl::exception(sycl::errc::invalid,
                            "Work item is not a valid coordination scope!");
    }
  } else if (is_gpu() && getBackend() == backend::ext_oneapi_level_zero) {
    switch (coordination_scope) {
    case execution_scope::root_group:
    case execution_scope::work_group:
      return forward_progress_guarantee::concurrent;
    case execution_scope::sub_group:
      return forward_progress_guarantee::weakly_parallel;
    default:
      throw sycl::exception(sycl::errc::invalid,
                            "Work item is not a valid coordination scope!");
    }
  }
  return forward_progress_guarantee::weakly_parallel;
}
 
} // namespace detail
} // namespace _V1
} // namespace sycl