commit 8216ca85037be9f4d5c20540522a22a4a93b660e
Author: Tobias Burnus <tobias@codesourcery.com>
Date:   Fri Jun 16 17:21:59 2023 +0200

    libgomp: Fix OMP_TARGET_OFFLOAD=mandatory
    
    It turned out that gomp_init_targets_once() was not run when directly
    calling 'omp target' or 'omp target (enter/exit) data' causing an
    abort with OMP_TARGET_OFFLOAD=mandatory wrongly claiming that no
    device is available. It was called a tiny bit later but few lines too
    late for updating the default-device-var.
    
    libgomp/ChangeLog:
    
            * target.c (resolve_device): Call gomp_get_num_devices early to ensure
            gomp_init_targets_once was called before using default-device-var.
            * testsuite/libgomp.c/target-55.c: New test.
            * testsuite/libgomp.c/target-55a.c: New test.
---
 libgomp/target.c                         | 10 +++++++---
 libgomp/testsuite/libgomp.c/target-55.c  | 20 ++++++++++++++++++++
 libgomp/testsuite/libgomp.c/target-55a.c | 23 +++++++++++++++++++++++
 3 files changed, 50 insertions(+), 3 deletions(-)

diff --git a/libgomp/target.c b/libgomp/target.c
index e39ef8f6e82..b6a7214ab4f 100644
--- a/libgomp/target.c
+++ b/libgomp/target.c
@@ -138,6 +138,10 @@ gomp_get_num_devices (void)
 static struct gomp_device_descr *
 resolve_device (int device_id, bool remapped)
 {
+  /* Get number of devices and thus ensure that 'gomp_init_targets_once' was
+     called, which must be done before using default_device_var.  */
+  int num_devices = gomp_get_num_devices ();
+
   if (remapped && device_id == GOMP_DEVICE_ICV)
     {
       struct gomp_task_icv *icv = gomp_icv (false);
@@ -151,7 +155,7 @@ resolve_device (int device_id, bool remapped)
 				 : omp_initial_device))
 	return NULL;
       if (gomp_target_offload_var == GOMP_TARGET_OFFLOAD_MANDATORY
-	  && gomp_get_num_devices () == 0)
+	  && num_devices == 0)
 	gomp_fatal ("OMP_TARGET_OFFLOAD is set to MANDATORY, "
 		    "but only the host device is available");
       else if (device_id == omp_invalid_device)
@@ -162,10 +166,10 @@ resolve_device (int device_id, bool remapped)
 
       return NULL;
     }
-  else if (device_id >= gomp_get_num_devices ())
+  else if (device_id >= num_devices)
     {
       if (gomp_target_offload_var == GOMP_TARGET_OFFLOAD_MANDATORY
-	  && device_id != num_devices_openmp)
+	  && device_id != num_devices)
 	gomp_fatal ("OMP_TARGET_OFFLOAD is set to MANDATORY, "
 		    "but device not found");
 
diff --git a/libgomp/testsuite/libgomp.c/target-55.c b/libgomp/testsuite/libgomp.c/target-55.c
new file mode 100644
index 00000000000..1314b3c6963
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c/target-55.c
@@ -0,0 +1,20 @@
+/* { dg-do run { target { offload_device } } } */
+/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "mandatory" } */
+
+/* Should pass - see target-55a.c for !offload_device */
+
+/* Check OMP_TARGET_OFFLOAD - it shall run on systems with offloading
+   devices available and fail otherwise.  Note that this did always
+   fail - as the device handling wasn't initialized before doing the
+   mandatory checking.  */
+
+int
+main ()
+{
+  int x = 1;
+  #pragma omp target map(tofrom: x)
+    x = 5;
+  if (x != 5)
+    __builtin_abort ();
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c/target-55a.c b/libgomp/testsuite/libgomp.c/target-55a.c
new file mode 100644
index 00000000000..53978c3f405
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c/target-55a.c
@@ -0,0 +1,23 @@
+/* { dg-do run { target { ! offload_device } } } */
+/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "mandatory" } */
+
+/* Should fail - see target-55a.c for offload_device */
+
+/* { dg-shouldfail "omp_invalid_device" } */
+/* { dg-output ".*libgomp: OMP_TARGET_OFFLOAD is set to MANDATORY, but only the host device is available.*" } */
+
+/* Check OMP_TARGET_OFFLOAD - it shall run on systems with offloading
+   devices available and fail otherwise.  Note that this did always
+   fail - as the device handling wasn't initialized before doing the
+   mandatory checking.  */
+
+int
+main ()
+{
+  int x = 1;
+  #pragma omp target map(tofrom: x)
+    x = 5;
+  if (x != 5)
+    __builtin_abort ();
+  return 0;
+}