Skip to content

Commit 3546ab8

Browse files
dm-vodopyanovdm-vodopyanov
authored
[SYCL][NFC] Refactor device architecture data structures (#12113)
The goal of changing macros to constexpr array of pairs is to have an ability to better maintain this data in the internal project.
1 parent 6cf1ae0 commit 3546ab8

File tree

1 file changed

+84
-80
lines changed

1 file changed

+84
-80
lines changed

‎sycl/source/detail/device_info.hpp

Lines changed: 84 additions & 80 deletions
Original file line numberDiff line numberDiff line change
@@ -557,94 +557,92 @@ struct get_device_info_impl<range<Dimensions>,
557557
}
558558
};
559559

560-
// This macro is only for AMD and NVIDIA GPU architectures
561-
#define NVIDIA_AMD_ARCHES(X) \
562-
X("5.0", oneapi_exp_arch::nvidia_gpu_sm_50) \
563-
X("5.2", oneapi_exp_arch::nvidia_gpu_sm_52) \
564-
X("5.3", oneapi_exp_arch::nvidia_gpu_sm_53) \
565-
X("6.0", oneapi_exp_arch::nvidia_gpu_sm_60) \
566-
X("6.1", oneapi_exp_arch::nvidia_gpu_sm_61) \
567-
X("6.2", oneapi_exp_arch::nvidia_gpu_sm_62) \
568-
X("7.0", oneapi_exp_arch::nvidia_gpu_sm_70) \
569-
X("7.2", oneapi_exp_arch::nvidia_gpu_sm_72) \
570-
X("7.5", oneapi_exp_arch::nvidia_gpu_sm_75) \
571-
X("8.0", oneapi_exp_arch::nvidia_gpu_sm_80) \
572-
X("8.6", oneapi_exp_arch::nvidia_gpu_sm_86) \
573-
X("8.7", oneapi_exp_arch::nvidia_gpu_sm_87) \
574-
X("8.9", oneapi_exp_arch::nvidia_gpu_sm_89) \
575-
X("9.0", oneapi_exp_arch::nvidia_gpu_sm_90) \
576-
X("gfx701", oneapi_exp_arch::amd_gpu_gfx701) \
577-
X("gfx702", oneapi_exp_arch::amd_gpu_gfx702) \
578-
X("gfx801", oneapi_exp_arch::amd_gpu_gfx801) \
579-
X("gfx802", oneapi_exp_arch::amd_gpu_gfx802) \
580-
X("gfx803", oneapi_exp_arch::amd_gpu_gfx803) \
581-
X("gfx805", oneapi_exp_arch::amd_gpu_gfx805) \
582-
X("gfx810", oneapi_exp_arch::amd_gpu_gfx810) \
583-
X("gfx900", oneapi_exp_arch::amd_gpu_gfx900) \
584-
X("gfx902", oneapi_exp_arch::amd_gpu_gfx902) \
585-
X("gfx904", oneapi_exp_arch::amd_gpu_gfx904) \
586-
X("gfx906", oneapi_exp_arch::amd_gpu_gfx906) \
587-
X("gfx908", oneapi_exp_arch::amd_gpu_gfx908) \
588-
X("gfx90a", oneapi_exp_arch::amd_gpu_gfx90a) \
589-
X("gfx1010", oneapi_exp_arch::amd_gpu_gfx1010) \
590-
X("gfx1011", oneapi_exp_arch::amd_gpu_gfx1011) \
591-
X("gfx1012", oneapi_exp_arch::amd_gpu_gfx1012) \
592-
X("gfx1013", oneapi_exp_arch::amd_gpu_gfx1013) \
593-
X("gfx1030", oneapi_exp_arch::amd_gpu_gfx1030) \
594-
X("gfx1031", oneapi_exp_arch::amd_gpu_gfx1031) \
595-
X("gfx1032", oneapi_exp_arch::amd_gpu_gfx1032) \
596-
X("gfx1034", oneapi_exp_arch::amd_gpu_gfx1034)
597-
598-
// This macro is only for Intel GPU architectures
599-
#define INTEL_GPU_ARCHES(X) \
600-
X(0x02000000, oneapi_exp_arch::intel_gpu_bdw) \
601-
X(0x02400009, oneapi_exp_arch::intel_gpu_skl) \
602-
X(0x02404009, oneapi_exp_arch::intel_gpu_kbl) \
603-
X(0x02408009, oneapi_exp_arch::intel_gpu_cfl) \
604-
X(0x0240c000, oneapi_exp_arch::intel_gpu_apl) \
605-
X(0x02410000, oneapi_exp_arch::intel_gpu_glk) \
606-
X(0x02414000, oneapi_exp_arch::intel_gpu_whl) \
607-
X(0x02418000, oneapi_exp_arch::intel_gpu_aml) \
608-
X(0x0241c000, oneapi_exp_arch::intel_gpu_cml) \
609-
X(0x02c00000, oneapi_exp_arch::intel_gpu_icllp) \
610-
X(0x02c08000, oneapi_exp_arch::intel_gpu_ehl) \
611-
X(0x03000000, oneapi_exp_arch::intel_gpu_tgllp) \
612-
X(0x03004000, oneapi_exp_arch::intel_gpu_rkl) \
613-
X(0x03008000, oneapi_exp_arch::intel_gpu_adl_s) \
614-
X(0x0300c000, oneapi_exp_arch::intel_gpu_adl_p) \
615-
X(0x03010000, oneapi_exp_arch::intel_gpu_adl_n) \
616-
X(0x03028000, oneapi_exp_arch::intel_gpu_dg1) \
617-
X(0x030dc008, oneapi_exp_arch::intel_gpu_acm_g10) \
618-
X(0x030e0005, oneapi_exp_arch::intel_gpu_acm_g11) \
619-
X(0x030e4000, oneapi_exp_arch::intel_gpu_acm_g12) \
620-
X(0x030f0007, oneapi_exp_arch::intel_gpu_pvc)
621-
622-
// This macro is only for Intel CPU architectures
623-
// TODO: extend the macro with other CPU architectures when they will be added
624-
// to ext_oneapi_device_architecture
625-
#define INTEL_CPU_ARCHES(X) \
626-
X(8, oneapi_exp_arch::intel_cpu_spr) \
627-
X(9, oneapi_exp_arch::intel_cpu_gnr)
628-
629-
#define CMP_NVIDIA_AMD(s, i) \
630-
if (strcmp(s, arch) == 0) \
631-
return i;
632-
633-
#define CMP_INTEL(p, i) \
634-
if (p == arch) \
635-
return i;
560+
using oneapi_exp_arch = sycl::ext::oneapi::experimental::architecture;
561+
562+
// Only for NVIDIA and AMD GPU architectures
563+
constexpr std::pair<const char *, oneapi_exp_arch> NvidiaAmdGPUArchitectures[] =
564+
{
565+
{"5.0", oneapi_exp_arch::nvidia_gpu_sm_50},
566+
{"5.2", oneapi_exp_arch::nvidia_gpu_sm_52},
567+
{"5.3", oneapi_exp_arch::nvidia_gpu_sm_53},
568+
{"6.0", oneapi_exp_arch::nvidia_gpu_sm_60},
569+
{"6.1", oneapi_exp_arch::nvidia_gpu_sm_61},
570+
{"6.2", oneapi_exp_arch::nvidia_gpu_sm_62},
571+
{"7.0", oneapi_exp_arch::nvidia_gpu_sm_70},
572+
{"7.2", oneapi_exp_arch::nvidia_gpu_sm_72},
573+
{"7.5", oneapi_exp_arch::nvidia_gpu_sm_75},
574+
{"8.0", oneapi_exp_arch::nvidia_gpu_sm_80},
575+
{"8.6", oneapi_exp_arch::nvidia_gpu_sm_86},
576+
{"8.7", oneapi_exp_arch::nvidia_gpu_sm_87},
577+
{"8.9", oneapi_exp_arch::nvidia_gpu_sm_89},
578+
{"9.0", oneapi_exp_arch::nvidia_gpu_sm_90},
579+
{"gfx701", oneapi_exp_arch::amd_gpu_gfx701},
580+
{"gfx702", oneapi_exp_arch::amd_gpu_gfx702},
581+
{"gfx801", oneapi_exp_arch::amd_gpu_gfx801},
582+
{"gfx802", oneapi_exp_arch::amd_gpu_gfx802},
583+
{"gfx803", oneapi_exp_arch::amd_gpu_gfx803},
584+
{"gfx805", oneapi_exp_arch::amd_gpu_gfx805},
585+
{"gfx810", oneapi_exp_arch::amd_gpu_gfx810},
586+
{"gfx900", oneapi_exp_arch::amd_gpu_gfx900},
587+
{"gfx902", oneapi_exp_arch::amd_gpu_gfx902},
588+
{"gfx904", oneapi_exp_arch::amd_gpu_gfx904},
589+
{"gfx906", oneapi_exp_arch::amd_gpu_gfx906},
590+
{"gfx908", oneapi_exp_arch::amd_gpu_gfx908},
591+
{"gfx90a", oneapi_exp_arch::amd_gpu_gfx90a},
592+
{"gfx1010", oneapi_exp_arch::amd_gpu_gfx1010},
593+
{"gfx1011", oneapi_exp_arch::amd_gpu_gfx1011},
594+
{"gfx1012", oneapi_exp_arch::amd_gpu_gfx1012},
595+
{"gfx1013", oneapi_exp_arch::amd_gpu_gfx1013},
596+
{"gfx1030", oneapi_exp_arch::amd_gpu_gfx1030},
597+
{"gfx1031", oneapi_exp_arch::amd_gpu_gfx1031},
598+
{"gfx1032", oneapi_exp_arch::amd_gpu_gfx1032},
599+
{"gfx1034", oneapi_exp_arch::amd_gpu_gfx1034},
600+
};
601+
602+
// Only for Intel GPU architectures
603+
constexpr std::pair<const int, oneapi_exp_arch> IntelGPUArchitectures[] = {
604+
{0x02000000, oneapi_exp_arch::intel_gpu_bdw},
605+
{0x02400009, oneapi_exp_arch::intel_gpu_skl},
606+
{0x02404009, oneapi_exp_arch::intel_gpu_kbl},
607+
{0x02408009, oneapi_exp_arch::intel_gpu_cfl},
608+
{0x0240c000, oneapi_exp_arch::intel_gpu_apl},
609+
{0x02410000, oneapi_exp_arch::intel_gpu_glk},
610+
{0x02414000, oneapi_exp_arch::intel_gpu_whl},
611+
{0x02418000, oneapi_exp_arch::intel_gpu_aml},
612+
{0x0241c000, oneapi_exp_arch::intel_gpu_cml},
613+
{0x02c00000, oneapi_exp_arch::intel_gpu_icllp},
614+
{0x02c08000, oneapi_exp_arch::intel_gpu_ehl},
615+
{0x03000000, oneapi_exp_arch::intel_gpu_tgllp},
616+
{0x03004000, oneapi_exp_arch::intel_gpu_rkl},
617+
{0x03008000, oneapi_exp_arch::intel_gpu_adl_s},
618+
{0x0300c000, oneapi_exp_arch::intel_gpu_adl_p},
619+
{0x03010000, oneapi_exp_arch::intel_gpu_adl_n},
620+
{0x03028000, oneapi_exp_arch::intel_gpu_dg1},
621+
{0x030dc008, oneapi_exp_arch::intel_gpu_acm_g10},
622+
{0x030e0005, oneapi_exp_arch::intel_gpu_acm_g11},
623+
{0x030e4000, oneapi_exp_arch::intel_gpu_acm_g12},
624+
{0x030f0007, oneapi_exp_arch::intel_gpu_pvc},
625+
};
626+
627+
// Only for Intel CPU architectures
628+
constexpr std::pair<const int, oneapi_exp_arch> IntelCPUArchitectures[] = {
629+
{8, oneapi_exp_arch::intel_cpu_spr},
630+
{9, oneapi_exp_arch::intel_cpu_gnr},
631+
};
636632

637633
template <>
638634
struct get_device_info_impl<
639635
ext::oneapi::experimental::architecture,
640636
ext::oneapi::experimental::info::device::architecture> {
641637
static ext::oneapi::experimental::architecture get(const DeviceImplPtr &Dev) {
642-
using oneapi_exp_arch = sycl::ext::oneapi::experimental::architecture;
643638
backend CurrentBackend = Dev->getBackend();
644639
if (Dev->is_gpu() && (backend::ext_oneapi_level_zero == CurrentBackend ||
645640
backend::opencl == CurrentBackend)) {
646641
auto MapArchIDToArchName = [](const int arch) {
647-
INTEL_GPU_ARCHES(CMP_INTEL);
642+
for (const auto &Item : IntelGPUArchitectures) {
643+
if (Item.first == arch)
644+
return Item.second;
645+
}
648646
throw sycl::exception(
649647
make_error_code(errc::runtime),
650648
"The current device architecture is not supported by "
@@ -660,7 +658,10 @@ struct get_device_info_impl<
660658
} else if (Dev->is_gpu() && (backend::ext_oneapi_cuda == CurrentBackend ||
661659
backend::ext_oneapi_hip == CurrentBackend)) {
662660
auto MapArchIDToArchName = [](const char *arch) {
663-
NVIDIA_AMD_ARCHES(CMP_NVIDIA_AMD);
661+
for (const auto &Item : NvidiaAmdGPUArchitectures) {
662+
if (std::string_view(Item.first) == arch)
663+
return Item.second;
664+
}
664665
throw sycl::exception(
665666
make_error_code(errc::runtime),
666667
"The current device architecture is not supported by "
@@ -680,7 +681,10 @@ struct get_device_info_impl<
680681
return MapArchIDToArchName(DeviceArchSubstr.data());
681682
} else if (Dev->is_cpu() && backend::opencl == CurrentBackend) {
682683
auto MapArchIDToArchName = [](const int arch) {
683-
INTEL_CPU_ARCHES(CMP_INTEL);
684+
for (const auto &Item : IntelCPUArchitectures) {
685+
if (Item.first == arch)
686+
return Item.second;
687+
}
684688
return sycl::ext::oneapi::experimental::architecture::x86_64;
685689
};
686690
uint32_t DeviceIp;

0 commit comments

Comments
 (0)