[llvm] [openmp] Reland: [OpenMP] Implement omp_get_uid_from_device() / omp_get_device_from_uid() (PR #168554)
Robert Imschweiler via llvm-commits
llvm-commits at lists.llvm.org
Tue Nov 18 07:52:35 PST 2025
https://github.com/ro-i created https://github.com/llvm/llvm-project/pull/168554
Reland https://github.com/llvm/llvm-project/pull/164392 with additional fix for fortran declarations (see second commit).
(@CatherineMoore fyi)
>From d070b75c4456c8dde94a5e54386c23217d893aab Mon Sep 17 00:00:00 2001
From: Robert Imschweiler <robert.imschweiler at amd.com>
Date: Mon, 20 Oct 2025 11:39:19 -0500
Subject: [PATCH 1/2] [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 | 58 ++++++++++++++++
offload/libomptarget/exports | 2 +
offload/test/api/omp_device_uid.c | 76 +++++++++++++++++++++
openmp/device/include/DeviceTypes.h | 3 +
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 | 29 +++++++-
openmp/runtime/src/kmp_ftn_os.h | 8 +++
openmp/runtime/test/api/omp_device_uid.c | 77 ++++++++++++++++++++++
15 files changed, 310 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 768ca46a9bed0..d92c7e450c677 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 89aa468689eaf..78e0d855c11e0 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 dd83a3ccd08e6..6e85e5764449c 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,
@@ -68,6 +70,62 @@ EXTERN int omp_get_device_num(void) {
return HostDevice;
}
+static inline bool is_initial_device_uid(const char *DeviceUid) {
+ return strcmp(DeviceUid, GenericPluginTy::getHostDeviceUid()) == 0;
+}
+
+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 (is_initial_device_uid(DeviceUid)) {
+ DP("Call to omp_get_device_from_uid returning initial 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 initial device UID\n");
+ return GenericPluginTy::getHostDeviceUid();
+ }
+
+ auto DeviceOrErr = PM->getDevice(DeviceNum);
+ if (!DeviceOrErr)
+ FATAL_MESSAGE(DeviceNum, "%s", toString(DeviceOrErr.takeError()).c_str());
+
+ const char *Uid =
+ DeviceOrErr->RTL->getDevice(DeviceOrErr->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 910a5b6c827a7..2ebc23e3cf60a 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..2a41d8d04ef8a
--- /dev/null
+++ b/offload/test/api/omp_device_uid.c
@@ -0,0 +1,76 @@
+// RUN: %libomptarget-compile-run-and-check-generic
+
+#include <omp.h>
+#include <stdio.h>
+#include <string.h>
+
+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 != 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 = omp_get_uid_from_device(device_num);
+
+ // omp_get_device_from_uid() in the device runtime is a dummy function
+ // returning omp_invalid_device.
+ int device_num_from_uid = omp_get_device_from_uid(device_uid);
+
+ // 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 ? device_num_from_uid != device_num
+ : 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);
+ 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..213ccfe58b4fb 100644
--- a/openmp/device/include/DeviceTypes.h
+++ b/openmp/device/include/DeviceTypes.h
@@ -21,6 +21,9 @@ 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 {
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..49c56d2b9a769 100644
--- a/openmp/runtime/src/kmp_ftn_entry.h
+++ b/openmp/runtime/src/kmp_ftn_entry.h
@@ -1543,13 +1543,38 @@ 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) {
+#if KMP_OS_DARWIN || KMP_OS_WASI || defined(KMP_STUB)
+ return nullptr;
+#else
+ const char *(*fptr)(int);
+ if ((*(void **)(&fptr) = KMP_DLSYM_NEXT("omp_get_uid_from_device")))
+ return (*fptr)(device_num);
+ // Returns the same string as used by libomptarget
+ return "HOST";
+#endif
+}
+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) {
+#if KMP_OS_DARWIN || KMP_OS_WASI || defined(KMP_STUB)
+ return omp_invalid_device;
+#else
+ int (*fptr)(const char *);
+ if ((*(void **)(&fptr) = KMP_DLSYM_NEXT("omp_get_device_from_uid")))
+ return (*fptr)(device_uid);
+ return KMP_EXPAND_NAME(FTN_GET_INITIAL_DEVICE)();
+#endif
+}
// 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,
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..40a1cbb644c7b
--- /dev/null
+++ b/openmp/runtime/test/api/omp_device_uid.c
@@ -0,0 +1,77 @@
+// RUN: %libomp-compile-and-run 2>&1 | FileCheck %s
+// Linking fails for icc 18
+// UNSUPPORTED: icc-18
+
+#include <omp_testsuite.h>
+#include <string.h>
+
+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 != 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 = omp_get_uid_from_device(device_num);
+
+ // omp_get_device_from_uid() in the device runtime is a dummy function
+ // returning omp_invalid_device.
+ int device_num_from_uid = omp_get_device_from_uid(device_uid);
+
+ // 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 ? device_num_from_uid != device_num
+ : 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);
+ 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 31b9b069f26f4f586acb7d941e63056eb7bf06ff Mon Sep 17 00:00:00 2001
From: Robert Imschweiler <robert.imschweiler at amd.com>
Date: Tue, 18 Nov 2025 09:48:52 -0600
Subject: [PATCH 2/2] fix fortran declarations
---
openmp/runtime/src/include/omp_lib.F90.var | 6 ++++--
openmp/runtime/src/include/omp_lib.h.var | 6 ++++--
2 files changed, 8 insertions(+), 4 deletions(-)
diff --git a/openmp/runtime/src/include/omp_lib.F90.var b/openmp/runtime/src/include/omp_lib.F90.var
index 159b42ab5b5cc..fa129ce5a9e9a 100644
--- a/openmp/runtime/src/include/omp_lib.F90.var
+++ b/openmp/runtime/src/include/omp_lib.F90.var
@@ -420,14 +420,16 @@
end function omp_get_device_num
function omp_get_uid_from_device(device_num) bind(c)
+ use, intrinsic :: iso_c_binding, only: c_ptr
use omp_lib_kinds
integer (kind=omp_integer_kind), value :: device_num
- character (len=*) omp_get_uid_from_device
+ type(c_ptr) omp_get_uid_from_device
end function omp_get_uid_from_device
function omp_get_device_from_uid(device_uid) bind(c)
+ use, intrinsic :: iso_c_binding, only: c_ptr
use omp_lib_kinds
- character (len=*), value :: device_uid
+ type(c_ptr), value :: device_uid
integer (kind=omp_integer_kind) omp_get_device_from_uid
end function omp_get_device_from_uid
diff --git a/openmp/runtime/src/include/omp_lib.h.var b/openmp/runtime/src/include/omp_lib.h.var
index 468eb03e99ef1..238cd5b77dc5b 100644
--- a/openmp/runtime/src/include/omp_lib.h.var
+++ b/openmp/runtime/src/include/omp_lib.h.var
@@ -491,13 +491,15 @@
function omp_get_uid_from_device(device_num) bind(c)
import
+ import :: c_ptr
integer (kind=omp_integer_kind), value :: device_num
- character (len=*) omp_get_uid_from_device
+ type(c_ptr) 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
+ import :: c_ptr
+ type(c_ptr), value :: device_uid
integer (kind=omp_integer_kind) omp_get_device_from_uid
end function omp_get_device_from_uid
More information about the llvm-commits
mailing list