https://github.com/ro-i updated https://github.com/llvm/llvm-project/pull/164392
>From 58e3ea4b9d9433f6e2662f2231915b2ac7be6ea6 Mon Sep 17 00:00:00 2001 From: Robert Imschweiler <[email protected]> Date: Mon, 20 Oct 2025 11:39:19 -0500 Subject: [PATCH 1/3] [OpenMP] Implement omp_get_uid_from_device() / omp_get_device_from_uid() Use the implementation in libomptarget. If libomptarget is not available, always return the UID / device number of the host / the initial device. --- offload/include/OpenMP/omp.h | 7 ++ offload/include/omptarget.h | 2 + offload/libomptarget/OpenMP/API.cpp | 55 +++++++++++++ offload/libomptarget/exports | 2 + offload/test/api/omp_device_uid.c | 88 +++++++++++++++++++++ openmp/device/include/DeviceTypes.h | 2 + openmp/device/include/Interface.h | 4 + openmp/device/src/State.cpp | 6 ++ openmp/runtime/src/dllexports | 2 + openmp/runtime/src/include/omp.h.var | 5 ++ openmp/runtime/src/include/omp_lib.F90.var | 14 ++++ openmp/runtime/src/include/omp_lib.h.var | 19 +++++ openmp/runtime/src/kmp_ftn_entry.h | 16 +++- openmp/runtime/src/kmp_ftn_os.h | 8 ++ openmp/runtime/test/api/omp_device_uid.c | 89 ++++++++++++++++++++++ 15 files changed, 317 insertions(+), 2 deletions(-) create mode 100644 offload/test/api/omp_device_uid.c create mode 100644 openmp/runtime/test/api/omp_device_uid.c diff --git a/offload/include/OpenMP/omp.h b/offload/include/OpenMP/omp.h index 49d9f1fa75c20..a42724f87cf3a 100644 --- a/offload/include/OpenMP/omp.h +++ b/offload/include/OpenMP/omp.h @@ -30,6 +30,13 @@ extern "C" { +/// Definitions +///{ + +#define omp_invalid_device -2 + +///} + /// Type declarations ///{ diff --git a/offload/include/omptarget.h b/offload/include/omptarget.h index 794b79e07674e..332b0d9f5cf27 100644 --- a/offload/include/omptarget.h +++ b/offload/include/omptarget.h @@ -274,6 +274,8 @@ extern "C" { void ompx_dump_mapping_tables(void); int omp_get_num_devices(void); int omp_get_device_num(void); +int omp_get_device_from_uid(const char *DeviceUid); +const char *omp_get_uid_from_device(int DeviceNum); int omp_get_initial_device(void); void *omp_target_alloc(size_t Size, int DeviceNum); void omp_target_free(void *DevicePtr, int DeviceNum); diff --git a/offload/libomptarget/OpenMP/API.cpp b/offload/libomptarget/OpenMP/API.cpp index b0f0573833713..cc4b607ccee6d 100644 --- a/offload/libomptarget/OpenMP/API.cpp +++ b/offload/libomptarget/OpenMP/API.cpp @@ -40,6 +40,8 @@ EXTERN void ompx_dump_mapping_tables() { using namespace llvm::omp::target::ompt; #endif +using GenericDeviceTy = llvm::omp::target::plugin::GenericDeviceTy; + void *targetAllocExplicit(size_t Size, int DeviceNum, int Kind, const char *Name); void targetFreeExplicit(void *DevicePtr, int DeviceNum, int Kind, @@ -91,6 +93,59 @@ EXTERN int omp_get_device_num(void) { return HostDevice; } +EXTERN int omp_get_device_from_uid(const char *DeviceUid) { + TIMESCOPE(); + OMPT_IF_BUILT(ReturnAddressSetterRAII RA(__builtin_return_address(0))); + + if (!DeviceUid) { + DP("Call to omp_get_device_from_uid returning omp_invalid_device\n"); + return omp_invalid_device; + } + if (strcmp(DeviceUid, GenericDeviceTy::getHostDeviceUid()) == 0) { + DP("Call to omp_get_device_from_uid returning host device number %d\n", + omp_get_initial_device()); + return omp_get_initial_device(); + } + + int DeviceNum = omp_invalid_device; + + auto ExclusiveDevicesAccessor = PM->getExclusiveDevicesAccessor(); + for (const DeviceTy &Device : PM->devices(ExclusiveDevicesAccessor)) { + const char *Uid = Device.RTL->getDevice(Device.RTLDeviceID).getDeviceUid(); + if (Uid && strcmp(DeviceUid, Uid) == 0) { + DeviceNum = Device.DeviceID; + break; + } + } + + DP("Call to omp_get_device_from_uid returning %d\n", DeviceNum); + return DeviceNum; +} + +EXTERN const char *omp_get_uid_from_device(int DeviceNum) { + TIMESCOPE(); + OMPT_IF_BUILT(ReturnAddressSetterRAII RA(__builtin_return_address(0))); + + if (DeviceNum == omp_invalid_device) { + DP("Call to omp_get_uid_from_device returning nullptr\n"); + return nullptr; + } + if (DeviceNum == omp_get_initial_device()) { + DP("Call to omp_get_uid_from_device returning host device UID\n"); + return GenericDeviceTy::getHostDeviceUid(); + } + + llvm::Expected<DeviceTy &> Device = PM->getDevice(DeviceNum); + if (!Device) { + FATAL_MESSAGE(DeviceNum, "%s", toString(Device.takeError()).c_str()); + return nullptr; + } + + const char *Uid = Device->RTL->getDevice(Device->RTLDeviceID).getDeviceUid(); + DP("Call to omp_get_uid_from_device returning %s\n", Uid); + return Uid; +} + EXTERN int omp_get_initial_device(void) { TIMESCOPE(); OMPT_IF_BUILT(ReturnAddressSetterRAII RA(__builtin_return_address(0))); diff --git a/offload/libomptarget/exports b/offload/libomptarget/exports index 1374bfea81511..f8fffb085a5dc 100644 --- a/offload/libomptarget/exports +++ b/offload/libomptarget/exports @@ -40,6 +40,8 @@ VERS1.0 { omp_get_mapped_ptr; omp_get_num_devices; omp_get_device_num; + omp_get_device_from_uid; + omp_get_uid_from_device; omp_get_initial_device; omp_target_alloc; omp_target_free; diff --git a/offload/test/api/omp_device_uid.c b/offload/test/api/omp_device_uid.c new file mode 100644 index 0000000000000..e07da8e56da98 --- /dev/null +++ b/offload/test/api/omp_device_uid.c @@ -0,0 +1,88 @@ +// RUN: %libomptarget-compile-run-and-check-generic + +#include <omp.h> +#include <stdio.h> +#include <string.h> + +// Note that the device UIDs for the "fake" host devices used by libomptarget +// will always be the same as the UID for the initial device (since it *is* the +// same device). The other way round, the device number returned for this UID +// will always be the initial device. + +int is_host_device_uid(const char *device_uid) { + return strcmp(device_uid, + omp_get_uid_from_device(omp_get_initial_device())) == 0; +} + +int test_omp_device_uid(int device_num) { + const char *device_uid = omp_get_uid_from_device(device_num); + if (device_uid == NULL) { + printf("FAIL for device %d: omp_get_uid_from_device returned NULL\n", + device_num); + return 0; + } + + int device_num_from_uid = omp_get_device_from_uid(device_uid); + if (device_num_from_uid != (is_host_device_uid(device_uid) + ? omp_get_initial_device() + : device_num)) { + printf( + "FAIL for device %d: omp_get_device_from_uid returned %d (UID: %s)\n", + device_num, device_num_from_uid, device_uid); + return 0; + } + + if (device_num == omp_get_initial_device()) + return 1; + + int success = 1; + +// Note that the following code may be executed on the host if the host is the +// device +#pragma omp target map(tofrom : success) device(device_num) + { + int device_num = omp_get_device_num(); + + // omp_get_uid_from_device() in the device runtime is a dummy function + // returning NULL + const char *device_uid_target = omp_get_uid_from_device(device_num); + + // omp_get_device_from_uid() in the device runtime is a dummy function + // returning omp_invalid_device. + device_num_from_uid = omp_get_device_from_uid(device_uid_target); + + // Depending on whether we're executing on the device or the host, we either + // got NULL as the device UID or the correct device UID. Consequently, + // omp_get_device_from_uid() either returned omp_invalid_device or the + // correct device number (aka omp_get_initial_device()). + if (device_uid_target ? device_num_from_uid != omp_get_initial_device() + : device_num_from_uid != omp_invalid_device) { + printf("FAIL for device %d (target): omp_get_device_from_uid returned %d " + "(UID: %s)\n", + device_num, device_num_from_uid, device_uid_target); + success = 0; + } + } + + return success; +} + +int main() { + int num_devices = omp_get_num_devices(); + int num_failed = 0; + // (also test initial device aka num_devices) + for (int i = 0; i < num_devices + 1; i++) { + if (!test_omp_device_uid(i)) { + printf("FAIL for device %d\n", i); + num_failed++; + } + } + if (num_failed) { + printf("FAIL\n"); + return 1; + } + printf("PASS\n"); + return 0; +} + +// CHECK: PASS diff --git a/openmp/device/include/DeviceTypes.h b/openmp/device/include/DeviceTypes.h index 2e5d92380f040..b7f9c5fe7b7cd 100644 --- a/openmp/device/include/DeviceTypes.h +++ b/openmp/device/include/DeviceTypes.h @@ -21,6 +21,8 @@ template <typename T> using Constant = __gpu_constant T; template <typename T> using Local = __gpu_local T; template <typename T> using Global = __gpu_local T; +#define omp_invalid_device -2 + enum omp_proc_bind_t { omp_proc_bind_false = 0, omp_proc_bind_true = 1, diff --git a/openmp/device/include/Interface.h b/openmp/device/include/Interface.h index c4bfaaa2404b4..71c3b1fc06d40 100644 --- a/openmp/device/include/Interface.h +++ b/openmp/device/include/Interface.h @@ -130,6 +130,10 @@ int omp_get_num_devices(void); int omp_get_device_num(void); +int omp_get_device_from_uid(const char *DeviceUid); + +const char *omp_get_uid_from_device(int DeviceNum); + int omp_get_num_teams(void); int omp_get_team_num(); diff --git a/openmp/device/src/State.cpp b/openmp/device/src/State.cpp index 475395102f47b..8ccb9d2ca24ff 100644 --- a/openmp/device/src/State.cpp +++ b/openmp/device/src/State.cpp @@ -423,6 +423,12 @@ int omp_get_num_devices(void) { return config::getNumDevices(); } int omp_get_device_num(void) { return config::getDeviceNum(); } +int omp_get_device_from_uid(const char *DeviceUid) { + return omp_invalid_device; +} + +const char *omp_get_uid_from_device(int DeviceNum) { return nullptr; } + int omp_get_num_teams(void) { return mapping::getNumberOfBlocksInKernel(); } int omp_get_team_num() { return mapping::getBlockIdInKernel(); } diff --git a/openmp/runtime/src/dllexports b/openmp/runtime/src/dllexports index 3983dae80c9f5..00becd1a657fd 100644 --- a/openmp/runtime/src/dllexports +++ b/openmp/runtime/src/dllexports @@ -544,6 +544,8 @@ kmp_set_disp_num_buffers 890 omp_get_devices_all_allocator 819 omp_get_memspace_num_resources 820 omp_get_submemspace 821 + omp_get_device_from_uid 822 + omp_get_uid_from_device 823 %ifndef stub __kmpc_set_default_allocator __kmpc_get_default_allocator diff --git a/openmp/runtime/src/include/omp.h.var b/openmp/runtime/src/include/omp.h.var index 74f385feb3ea5..e98df731ad888 100644 --- a/openmp/runtime/src/include/omp.h.var +++ b/openmp/runtime/src/include/omp.h.var @@ -536,6 +536,11 @@ /* OpenMP 5.2 */ extern int __KAI_KMPC_CONVENTION omp_in_explicit_task(void); + #define omp_invalid_device -2 + + /* OpenMP 6.0 */ + extern int __KAI_KMPC_CONVENTION omp_get_device_from_uid(const char *DeviceUid); + extern const char * __KAI_KMPC_CONVENTION omp_get_uid_from_device(int DeviceNum); /* LLVM Extensions */ extern void *llvm_omp_target_dynamic_shared_alloc(void); diff --git a/openmp/runtime/src/include/omp_lib.F90.var b/openmp/runtime/src/include/omp_lib.F90.var index 90d7e49ebf549..159b42ab5b5cc 100644 --- a/openmp/runtime/src/include/omp_lib.F90.var +++ b/openmp/runtime/src/include/omp_lib.F90.var @@ -215,6 +215,8 @@ integer (kind=omp_interop_kind), parameter, public :: omp_interop_none = 0 + integer (kind=omp_integer_kind), parameter, public :: omp_invalid_device = -2 + interface ! *** @@ -417,6 +419,18 @@ integer (kind=omp_integer_kind) omp_get_device_num end function omp_get_device_num + function omp_get_uid_from_device(device_num) bind(c) + use omp_lib_kinds + integer (kind=omp_integer_kind), value :: device_num + character (len=*) omp_get_uid_from_device + end function omp_get_uid_from_device + + function omp_get_device_from_uid(device_uid) bind(c) + use omp_lib_kinds + character (len=*), value :: device_uid + integer (kind=omp_integer_kind) omp_get_device_from_uid + end function omp_get_device_from_uid + function omp_pause_resource(kind, device_num) bind(c) use omp_lib_kinds integer (kind=omp_pause_resource_kind), value :: kind diff --git a/openmp/runtime/src/include/omp_lib.h.var b/openmp/runtime/src/include/omp_lib.h.var index a50bb018c7cc3..468eb03e99ef1 100644 --- a/openmp/runtime/src/include/omp_lib.h.var +++ b/openmp/runtime/src/include/omp_lib.h.var @@ -291,6 +291,9 @@ integer(kind=omp_interop_kind)omp_interop_none parameter(omp_interop_none=0) + integer(kind=omp_integer_kind)omp_invalid_device + parameter(omp_invalid_device=-2) + interface ! *** @@ -486,6 +489,18 @@ integer (kind=omp_integer_kind) omp_get_device_num end function omp_get_device_num + function omp_get_uid_from_device(device_num) bind(c) + import + integer (kind=omp_integer_kind), value :: device_num + character (len=*) omp_get_uid_from_device + end function omp_get_uid_from_device + + function omp_get_device_from_uid(device_uid) bind(c) + import + character (len=*), value :: device_uid + integer (kind=omp_integer_kind) omp_get_device_from_uid + end function omp_get_device_from_uid + function omp_pause_resource(kind, device_num) bind(c) import integer (kind=omp_pause_resource_kind), value :: kind @@ -1159,6 +1174,8 @@ !DIR$ ATTRIBUTES OFFLOAD:MIC :: omp_get_initial_device !DIR$ ATTRIBUTES OFFLOAD:MIC :: omp_get_num_devices !DIR$ ATTRIBUTES OFFLOAD:MIC :: omp_get_device_num +!DIR$ ATTRIBUTES OFFLOAD:MIC :: omp_get_uid_from_device +!DIR$ ATTRIBUTES OFFLOAD:MIC :: omp_get_device_from_uid !DIR$ ATTRIBUTES OFFLOAD:MIC :: omp_pause_resource !DIR$ ATTRIBUTES OFFLOAD:MIC :: omp_pause_resource_all !DIR$ ATTRIBUTES OFFLOAD:MIC :: omp_get_supported_active_levels @@ -1242,6 +1259,8 @@ !$omp declare target(omp_get_initial_device ) !$omp declare target(omp_get_num_devices ) !$omp declare target(omp_get_device_num ) +!$omp declare target(omp_get_uid_from_device ) +!$omp declare target(omp_get_device_from_uid ) !$omp declare target(omp_pause_resource ) !$omp declare target(omp_pause_resource_all ) !$omp declare target(omp_get_supported_active_levels ) diff --git a/openmp/runtime/src/kmp_ftn_entry.h b/openmp/runtime/src/kmp_ftn_entry.h index 2b0063eb23a0a..95cd3965a518d 100644 --- a/openmp/runtime/src/kmp_ftn_entry.h +++ b/openmp/runtime/src/kmp_ftn_entry.h @@ -1543,14 +1543,26 @@ int FTN_STDCALL KMP_EXPAND_NAME(FTN_GET_MAX_TASK_PRIORITY)(void) { #endif } -// This function will be defined in libomptarget. When libomptarget is not -// loaded, we assume we are on the host and return KMP_HOST_DEVICE. +// These functions will be defined in libomptarget. When libomptarget is not +// loaded, we assume we are on the host. // Compiler/libomptarget will handle this if called inside target. int FTN_STDCALL FTN_GET_DEVICE_NUM(void) KMP_WEAK_ATTRIBUTE_EXTERNAL; int FTN_STDCALL FTN_GET_DEVICE_NUM(void) { return KMP_EXPAND_NAME(FTN_GET_INITIAL_DEVICE)(); } +const char *FTN_STDCALL FTN_GET_UID_FROM_DEVICE(int device_num) + KMP_WEAK_ATTRIBUTE_EXTERNAL; +const char *FTN_STDCALL FTN_GET_UID_FROM_DEVICE(int device_num) { + // Returns the same string as used by libomptarget + return "HOST"; +} +int FTN_STDCALL FTN_GET_DEVICE_FROM_UID(const char *device_uid) + KMP_WEAK_ATTRIBUTE_EXTERNAL; +int FTN_STDCALL FTN_GET_DEVICE_FROM_UID(const char *device_uid) { + return KMP_EXPAND_NAME(FTN_GET_INITIAL_DEVICE)(); +} + // Compiler will ensure that this is only called from host in sequential region int FTN_STDCALL KMP_EXPAND_NAME(FTN_PAUSE_RESOURCE)(kmp_pause_status_t kind, int device_num) { diff --git a/openmp/runtime/src/kmp_ftn_os.h b/openmp/runtime/src/kmp_ftn_os.h index ae0ed067235e5..c439a058f22b4 100644 --- a/openmp/runtime/src/kmp_ftn_os.h +++ b/openmp/runtime/src/kmp_ftn_os.h @@ -140,6 +140,8 @@ #define FTN_GET_MEMSPACE_NUM_RESOURCES omp_get_memspace_num_resources #define FTN_GET_SUBMEMSPACE omp_get_submemspace #define FTN_GET_DEVICE_NUM omp_get_device_num +#define FTN_GET_UID_FROM_DEVICE omp_get_uid_from_device +#define FTN_GET_DEVICE_FROM_UID omp_get_device_from_uid #define FTN_SET_AFFINITY_FORMAT omp_set_affinity_format #define FTN_GET_AFFINITY_FORMAT omp_get_affinity_format #define FTN_DISPLAY_AFFINITY omp_display_affinity @@ -289,6 +291,8 @@ #define FTN_ALLOC omp_alloc_ #define FTN_FREE omp_free_ #define FTN_GET_DEVICE_NUM omp_get_device_num_ +#define FTN_GET_UID_FROM_DEVICE omp_get_uid_from_device_ +#define FTN_GET_DEVICE_FROM_UID omp_get_device_from_uid_ #define FTN_SET_AFFINITY_FORMAT omp_set_affinity_format_ #define FTN_GET_AFFINITY_FORMAT omp_get_affinity_format_ #define FTN_DISPLAY_AFFINITY omp_display_affinity_ @@ -436,6 +440,8 @@ #define FTN_GET_MEMSPACE_NUM_RESOURCES OMP_GET_MEMSPACE_NUM_RESOURCES #define FTN_GET_SUBMEMSPACE OMP_GET_SUBMEMSPACE #define FTN_GET_DEVICE_NUM OMP_GET_DEVICE_NUM +#define FTN_GET_UID_FROM_DEVICE OMP_GET_UID_FROM_DEVICE +#define FTN_GET_DEVICE_FROM_UID OMP_GET_DEVICE_FROM_UID #define FTN_SET_AFFINITY_FORMAT OMP_SET_AFFINITY_FORMAT #define FTN_GET_AFFINITY_FORMAT OMP_GET_AFFINITY_FORMAT #define FTN_DISPLAY_AFFINITY OMP_DISPLAY_AFFINITY @@ -585,6 +591,8 @@ #define FTN_ALLOC OMP_ALLOC_ #define FTN_FREE OMP_FREE_ #define FTN_GET_DEVICE_NUM OMP_GET_DEVICE_NUM_ +#define FTN_GET_UID_FROM_DEVICE OMP_GET_UID_FROM_DEVICE_ +#define FTN_GET_DEVICE_FROM_UID OMP_GET_DEVICE_FROM_UID_ #define FTN_SET_AFFINITY_FORMAT OMP_SET_AFFINITY_FORMAT_ #define FTN_GET_AFFINITY_FORMAT OMP_GET_AFFINITY_FORMAT_ #define FTN_DISPLAY_AFFINITY OMP_DISPLAY_AFFINITY_ diff --git a/openmp/runtime/test/api/omp_device_uid.c b/openmp/runtime/test/api/omp_device_uid.c new file mode 100644 index 0000000000000..975dad73f5b18 --- /dev/null +++ b/openmp/runtime/test/api/omp_device_uid.c @@ -0,0 +1,89 @@ +// RUN: %libomp-compile-and-run 2>&1 | FileCheck %s +// Linking fails for icc 18 +// UNSUPPORTED: icc-18 + +#include <omp_testsuite.h> +#include <string.h> + +// Note that the device UIDs for the "fake" host devices used by libomptarget +// will always be the same as the UID for the initial device (since it *is* the +// same device). The other way round, the device number returned for this UID +// will always be the initial device. + +int is_host_device_uid(const char *device_uid) { + return strcmp(device_uid, + omp_get_uid_from_device(omp_get_initial_device())) == 0; +} + +int test_omp_device_uid(int device_num) { + const char *device_uid = omp_get_uid_from_device(device_num); + if (device_uid == NULL) { + printf("FAIL for device %d: omp_get_uid_from_device returned NULL\n", + device_num); + return 0; + } + + int device_num_from_uid = omp_get_device_from_uid(device_uid); + if (device_num_from_uid != (is_host_device_uid(device_uid) + ? omp_get_initial_device() + : device_num)) { + printf( + "FAIL for device %d: omp_get_device_from_uid returned %d (UID: %s)\n", + device_num, device_num_from_uid, device_uid); + return 0; + } + + if (device_num == omp_get_initial_device()) + return 1; + + int success = 1; + +// Note that the following code may be executed on the host if the host is the +// device +#pragma omp target map(tofrom : success) device(device_num) + { + int device_num = omp_get_device_num(); + + // omp_get_uid_from_device() in the device runtime is a dummy function + // returning NULL + const char *device_uid_target = omp_get_uid_from_device(device_num); + + // omp_get_device_from_uid() in the device runtime is a dummy function + // returning omp_invalid_device. + device_num_from_uid = omp_get_device_from_uid(device_uid_target); + + // Depending on whether we're executing on the device or the host, we either + // got NULL as the device UID or the correct device UID. Consequently, + // omp_get_device_from_uid() either returned omp_invalid_device or the + // correct device number (aka omp_get_initial_device()). + if (device_uid_target ? device_num_from_uid != omp_get_initial_device() + : device_num_from_uid != omp_invalid_device) { + printf("FAIL for device %d (target): omp_get_device_from_uid returned %d " + "(UID: %s)\n", + device_num, device_num_from_uid, device_uid_target); + success = 0; + } + } + + return success; +} + +int main() { + int num_devices = omp_get_num_devices(); + int num_failed = 0; + // (also test initial device aka num_devices) + for (int i = 0; i < num_devices + 1; i++) { + if (!test_omp_device_uid(i)) { + printf("FAIL for device %d\n", i); + num_failed++; + } + } + if (num_failed) { + printf("FAIL\n"); + return 1; + } + printf("PASS\n"); + return 0; +} + +// CHECK: PASS >From e6db60b2c2052fa222177e948ad2d44511955b3e Mon Sep 17 00:00:00 2001 From: Robert Imschweiler <[email protected]> Date: Tue, 21 Oct 2025 09:13:46 -0500 Subject: [PATCH 2/3] implement feedback --- offload/libomptarget/OpenMP/API.cpp | 6 +++++- openmp/device/include/DeviceTypes.h | 1 + openmp/runtime/src/kmp_ftn_entry.h | 1 - 3 files changed, 6 insertions(+), 2 deletions(-) diff --git a/offload/libomptarget/OpenMP/API.cpp b/offload/libomptarget/OpenMP/API.cpp index cc4b607ccee6d..eb83f9ff3c18f 100644 --- a/offload/libomptarget/OpenMP/API.cpp +++ b/offload/libomptarget/OpenMP/API.cpp @@ -93,6 +93,10 @@ EXTERN int omp_get_device_num(void) { return HostDevice; } +static inline bool is_host_device_uid(const char *DeviceUid) { + return strcmp(DeviceUid, GenericDeviceTy::getHostDeviceUid()) == 0; +} + EXTERN int omp_get_device_from_uid(const char *DeviceUid) { TIMESCOPE(); OMPT_IF_BUILT(ReturnAddressSetterRAII RA(__builtin_return_address(0))); @@ -101,7 +105,7 @@ EXTERN int omp_get_device_from_uid(const char *DeviceUid) { DP("Call to omp_get_device_from_uid returning omp_invalid_device\n"); return omp_invalid_device; } - if (strcmp(DeviceUid, GenericDeviceTy::getHostDeviceUid()) == 0) { + if (is_host_device_uid(DeviceUid)) { DP("Call to omp_get_device_from_uid returning host device number %d\n", omp_get_initial_device()); return omp_get_initial_device(); diff --git a/openmp/device/include/DeviceTypes.h b/openmp/device/include/DeviceTypes.h index b7f9c5fe7b7cd..213ccfe58b4fb 100644 --- a/openmp/device/include/DeviceTypes.h +++ b/openmp/device/include/DeviceTypes.h @@ -21,6 +21,7 @@ template <typename T> using Constant = __gpu_constant T; template <typename T> using Local = __gpu_local T; template <typename T> using Global = __gpu_local T; +// See definition in OpenMP (omp.h.var/omp_lib.(F90|h).var) #define omp_invalid_device -2 enum omp_proc_bind_t { diff --git a/openmp/runtime/src/kmp_ftn_entry.h b/openmp/runtime/src/kmp_ftn_entry.h index 95cd3965a518d..944ccdf241f70 100644 --- a/openmp/runtime/src/kmp_ftn_entry.h +++ b/openmp/runtime/src/kmp_ftn_entry.h @@ -1550,7 +1550,6 @@ int FTN_STDCALL FTN_GET_DEVICE_NUM(void) KMP_WEAK_ATTRIBUTE_EXTERNAL; int FTN_STDCALL FTN_GET_DEVICE_NUM(void) { return KMP_EXPAND_NAME(FTN_GET_INITIAL_DEVICE)(); } - const char *FTN_STDCALL FTN_GET_UID_FROM_DEVICE(int device_num) KMP_WEAK_ATTRIBUTE_EXTERNAL; const char *FTN_STDCALL FTN_GET_UID_FROM_DEVICE(int device_num) { >From a3171c2e309dc3e5582be484ccf9a1beb5eaa1ab Mon Sep 17 00:00:00 2001 From: Robert Imschweiler <[email protected]> Date: Tue, 21 Oct 2025 15:13:56 -0500 Subject: [PATCH 3/3] adapt to offload plugin device UID getter --- offload/libomptarget/OpenMP/API.cpp | 8 ++++---- 1 file changed, 4 insertions(+), 4 deletions(-) diff --git a/offload/libomptarget/OpenMP/API.cpp b/offload/libomptarget/OpenMP/API.cpp index eb83f9ff3c18f..871873cbf506e 100644 --- a/offload/libomptarget/OpenMP/API.cpp +++ b/offload/libomptarget/OpenMP/API.cpp @@ -94,7 +94,7 @@ EXTERN int omp_get_device_num(void) { } static inline bool is_host_device_uid(const char *DeviceUid) { - return strcmp(DeviceUid, GenericDeviceTy::getHostDeviceUid()) == 0; + return strcmp(DeviceUid, GenericPluginTy::getHostDeviceUid()) == 0; } EXTERN int omp_get_device_from_uid(const char *DeviceUid) { @@ -115,7 +115,7 @@ EXTERN int omp_get_device_from_uid(const char *DeviceUid) { auto ExclusiveDevicesAccessor = PM->getExclusiveDevicesAccessor(); for (const DeviceTy &Device : PM->devices(ExclusiveDevicesAccessor)) { - const char *Uid = Device.RTL->getDevice(Device.RTLDeviceID).getDeviceUid(); + const char *Uid = Device.RTL->getDeviceUid(Device.RTLDeviceID); if (Uid && strcmp(DeviceUid, Uid) == 0) { DeviceNum = Device.DeviceID; break; @@ -136,7 +136,7 @@ EXTERN const char *omp_get_uid_from_device(int DeviceNum) { } if (DeviceNum == omp_get_initial_device()) { DP("Call to omp_get_uid_from_device returning host device UID\n"); - return GenericDeviceTy::getHostDeviceUid(); + return GenericPluginTy::getHostDeviceUid(); } llvm::Expected<DeviceTy &> Device = PM->getDevice(DeviceNum); @@ -145,7 +145,7 @@ EXTERN const char *omp_get_uid_from_device(int DeviceNum) { return nullptr; } - const char *Uid = Device->RTL->getDevice(Device->RTLDeviceID).getDeviceUid(); + const char *Uid = Device->RTL->getDeviceUid(Device->RTLDeviceID); DP("Call to omp_get_uid_from_device returning %s\n", Uid); return Uid; } _______________________________________________ llvm-branch-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-branch-commits
