[gcn] Use 'radeon' for the environment variable 'ACC_DEVICE_TYPE'

..., per OpenACC 3.0, A.1.2. "AMD GPU Targets".

This complements commit 6687d13a87c42dddc7d1c7adade38d31ba0d1401 "Rename
acc_device_gcn to acc_device_radeon".

	libgomp/
	* oacc-init.c (get_openacc_name): Handle 'gcn'.
	* testsuite/lib/libgomp.exp
	(offload_target_to_openacc_device_type) [amdgcn*]: Return
	'radeon'.  Adjust all users.
	(check_effective_target_openacc_amdgcn_accel_present): Rename
	to...
	(check_effective_target_openacc_radeon_accel_present): ... this.
	Adjust all users.
	(check_effective_target_openacc_amdgcn_accel_selected): Rename to...
	(check_effective_target_openacc_radeon_accel_selected): ... this.
	Adjust all users.

diff --git a/libgomp/ChangeLog b/libgomp/ChangeLog
index 53bb8d23fa1524828354de5a5e1d5cac473cffce..cfe6e0653c92fb0c419662b22ff351458f1e7c19 100644 (file)
--- a/libgomp/ChangeLog
+++ b/libgomp/ChangeLog
@@ -1,5 +1,17 @@
 2020-04-29  Thomas Schwinge  <thomas@codesourcery.com>
 
+       * oacc-init.c (get_openacc_name): Handle 'gcn'.
+       * testsuite/lib/libgomp.exp
+       (offload_target_to_openacc_device_type) [amdgcn*]: Return
+       'radeon'.  Adjust all users.
+       (check_effective_target_openacc_amdgcn_accel_present): Rename
+       to...
+       (check_effective_target_openacc_radeon_accel_present): ... this.
+       Adjust all users.
+       (check_effective_target_openacc_amdgcn_accel_selected): Rename to...
+       (check_effective_target_openacc_radeon_accel_selected): ... this.
+       Adjust all users.
+
        * testsuite/libgomp.fortran/use_device_ptr-optional-2.f90: Add
        'dg-do run'.
 

diff --git a/libgomp/oacc-init.c b/libgomp/oacc-init.c
index ef12b4c16d014cb134f611c6d0f8666779df2c07..5d786a5a2e7c8ce26cc878c83ba332ca4a3a895e 100644 (file)
--- a/libgomp/oacc-init.c
+++ b/libgomp/oacc-init.c
@@ -99,7 +99,9 @@ unknown_device_type_error (acc_device_t invalid_type)
 static const char *
 get_openacc_name (const char *name)
 {
-  if (strcmp (name, "nvptx") == 0)
+  if (strcmp (name, "gcn") == 0)
+    return "radeon";
+  else if (strcmp (name, "nvptx") == 0)
     return "nvidia";
   else
     return name;

diff --git a/libgomp/testsuite/lib/libgomp.exp b/libgomp/testsuite/lib/libgomp.exp
index e7ce784314d915371b4771ad068d393264dd11f3..ee5f0e57b19038021f4efdc26ffd24ad9843337a 100644 (file)
--- a/libgomp/testsuite/lib/libgomp.exp
+++ b/libgomp/testsuite/lib/libgomp.exp
@@ -319,7 +319,7 @@ proc libgomp_option_proc { option } {
 proc offload_target_to_openacc_device_type { offload_target } {
     switch -glob $offload_target {
        amdgcn* {
-           return "gcn"
+           return "radeon"
        }
        disable {
            return "host"
@@ -483,10 +483,10 @@ proc check_effective_target_hsa_offloading_selected {} {
     }]
 }
 
-# Return 1 if at least one AMD GCN board is present.
+# Return 1 if at least one AMD GPU is accessible.
 
-proc check_effective_target_openacc_amdgcn_accel_present { } {
-    return [check_runtime openacc_amdgcn_accel_present {
+proc check_effective_target_openacc_radeon_accel_present { } {
+    return [check_runtime openacc_radeon_accel_present {
        #include <openacc.h>
        int main () {
        return !(acc_get_num_devices (acc_device_radeon) > 0);
@@ -494,11 +494,11 @@ proc check_effective_target_openacc_amdgcn_accel_present { } {
     } "" ]
 }
 
-# Return 1 if at least one AMD GCN board is present, and the AMD GCN device
-# type is selected by default.
+# Return 1 if at least one AMD GPU is accessible, and the OpenACC 'radeon'
+# device type is selected.
 
-proc check_effective_target_openacc_amdgcn_accel_selected { } {
-    if { ![check_effective_target_openacc_amdgcn_accel_present] } {
+proc check_effective_target_openacc_radeon_accel_selected { } {
+    if { ![check_effective_target_openacc_radeon_accel_present] } {
        return 0;
     }
     global offload_target

diff --git a/libgomp/testsuite/libgomp.oacc-c++/c++.exp b/libgomp/testsuite/libgomp.oacc-c++/c++.exp
index c06c2a097e38f9da33e65160857f640457731c9c..7200ec19c476cf5f6a369a34d02fea27bba83202 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c++/c++.exp
+++ b/libgomp/testsuite/libgomp.oacc-c++/c++.exp
@@ -88,15 +88,6 @@ if { $lang_test_file_found } {
                unsupported "$subdir $offload_target offloading"
                continue
            }
-           gcn {
-               if { ![check_effective_target_openacc_amdgcn_accel_present] } {
-                   # Don't bother; execution testing is going to FAIL.
-                   untested "$subdir $offload_target offloading: supported, but hardware not accessible"
-                   continue
-               }
-
-               set acc_mem_shared 0
-           }
            host {
                set acc_mem_shared 1
            }
@@ -115,6 +106,15 @@ if { $lang_test_file_found } {
 
                set acc_mem_shared 0
            }
+           radeon {
+               if { ![check_effective_target_openacc_radeon_accel_present] } {
+                   # Don't bother; execution testing is going to FAIL.
+                   untested "$subdir $offload_target offloading: supported, but hardware not accessible"
+                   continue
+               }
+
+               set acc_mem_shared 0
+           }
            default {
                error "Unknown OpenACC device type: $openacc_device_type (offload target: $offload_target)"
            }

diff --git a/libgomp/testsuite/libgomp.oacc-c++/firstprivate-mappings-1.C b/libgomp/testsuite/libgomp.oacc-c++/firstprivate-mappings-1.C
index 7b3e670073c9beddb89feb083a1f2a5977b08266..b046bf2912d6fc881d58fb534dc0bd5ef0affa22 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c++/firstprivate-mappings-1.C
+++ b/libgomp/testsuite/libgomp.oacc-c++/firstprivate-mappings-1.C
@@ -3,7 +3,7 @@
 /* PR middle-end/48591 */
 /* PR other/71064 */
 /* Set to 0 for offloading targets not supporting long double.  */
-#if defined(ACC_DEVICE_TYPE_nvidia) || defined(ACC_DEVICE_TYPE_gcn)
+#if defined(ACC_DEVICE_TYPE_nvidia) || defined(ACC_DEVICE_TYPE_radeon)
 # define DO_LONG_DOUBLE 0
 #else
 # define DO_LONG_DOUBLE 1

diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_get_property-gcn.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_get_property-gcn.c
index ce59264a60dc84e18356a7a39e76db53f8ad5439..4b1fb5e0e761d32ffbf9da60176fdb2bf3a1084c 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_get_property-gcn.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/acc_get_property-gcn.c
@@ -3,7 +3,7 @@
    those obtained through the HSA API. */
 /* { dg-additional-sources acc_get_property-aux.c } */
 /* { dg-additional-options "-ldl" } */
-/* { dg-do run { target openacc_amdgcn_accel_selected } } */
+/* { dg-do run { target openacc_radeon_accel_selected } } */
 
 #include <dlfcn.h>
 #include <stdint.h>

diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/asyncwait-nop-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/asyncwait-nop-1.c
index 754e015a280cc7bd94e38af1e0ba8c83652b7e2b..7496426c8fa3c5001b774a870ed0394514e2bbd5 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/asyncwait-nop-1.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/asyncwait-nop-1.c
@@ -26,7 +26,7 @@ main ()
   acc_device_t d;
 #if defined ACC_DEVICE_TYPE_nvidia
   d = acc_device_nvidia;
-#elif defined ACC_DEVICE_TYPE_gcn
+#elif defined ACC_DEVICE_TYPE_radeon
   d = acc_device_radeon;
 #elif defined ACC_DEVICE_TYPE_host
   d = acc_device_host;

diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/firstprivate-mappings-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/firstprivate-mappings-1.c
index 253f8bf0bd0267b39ab69c0da6d784cfb8b8b0f4..2cdd2d1525e785ac18498479737212ccab30c762 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/firstprivate-mappings-1.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/firstprivate-mappings-1.c
@@ -6,7 +6,7 @@
 /* PR middle-end/48591 */
 /* PR other/71064 */
 /* Set to 0 for offloading targets not supporting long double.  */
-#if defined(ACC_DEVICE_TYPE_nvidia) || defined(ACC_DEVICE_TYPE_gcn)
+#if defined(ACC_DEVICE_TYPE_nvidia) || defined(ACC_DEVICE_TYPE_radeon)
 # define DO_LONG_DOUBLE 0
 #else
 # define DO_LONG_DOUBLE 1

diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/function-not-offloaded.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/function-not-offloaded.c
index 517004a562d2a96cf2f84db7c8487d0afc7cc107..64f8ab812a63d6e1f70737b9be2e7e1e0f157e01 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/function-not-offloaded.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/function-not-offloaded.c
@@ -1,11 +1,11 @@
 /* { dg-do link } */
-/* { dg-excess-errors "lto1, mkoffload and lto-wrapper fatal errors" { target { openacc_nvidia_accel_selected || openacc_amdgcn_accel_selected } } } */
+/* { dg-excess-errors "lto1, mkoffload and lto-wrapper fatal errors" { target { openacc_nvidia_accel_selected || openacc_radeon_accel_selected } } } */
 
 int var;
 #pragma acc declare create (var)
 
 void __attribute__((noinline, noclone))
-foo () /* { dg-error "function 'foo' has been referenced in offloaded code but hasn't been marked to be included in the offloaded code" "" { target { openacc_nvidia_accel_selected || openacc_amdgcn_accel_selected } } } */
+foo () /* { dg-error "function 'foo' has been referenced in offloaded code but hasn't been marked to be included in the offloaded code" "" { target { openacc_nvidia_accel_selected || openacc_radeon_accel_selected } } } */
 {
   var++;
 }

diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-auto-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-auto-1.c
index 0c9ae957460103aaa8f283462b0bc66b4090256b..0273c2bddd74d4b675bd145d176f25d3fe843159 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-auto-1.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-auto-1.c
@@ -1,5 +1,5 @@
 /* AMD GCN does not use 32-lane vectors.
-   { dg-skip-if "unsuitable dimensions" { openacc_amdgcn_accel_selected } { "*" } { "" } } */
+   { dg-skip-if "unsuitable dimensions" { openacc_radeon_accel_selected } { "*" } { "" } } */
 
 /* { dg-additional-options "-fopenacc-dim=32" } */
 

diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-dim-default.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-dim-default.c
index 30f0539707f5d7de95cd916601770f79de878eac..ca7716466558c2eae164aa36adbce08fd2afbc30 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-dim-default.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-dim-default.c
@@ -128,7 +128,7 @@ int test_1 (int gp, int wp, int vp)
 
 int main ()
 {
-#ifdef ACC_DEVICE_TYPE_gcn
+#ifdef ACC_DEVICE_TYPE_radeon
   /* AMD GCN uses the autovectorizer for the vector dimension: the use
      of a function call in vector-partitioned code in this test is not
      currently supported.  */

diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-wv-2.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-wv-2.c
index 609f9f6a7da2302ce3f55af895b2e114aa20fdfd..9769ee7243085ddbbf82de86c57b71f14935fdd5 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-wv-2.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-wv-2.c
@@ -2,7 +2,7 @@
 #include <openacc.h>
 #include <gomp-constants.h>
 
-#ifdef ACC_DEVICE_TYPE_gcn
+#ifdef ACC_DEVICE_TYPE_radeon
 #define NUM_WORKERS 16
 #define NUM_VECTORS 1
 #else

diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/tile-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/tile-1.c
index c019fe55c7a2250d93d0662609e21416514fd123..5757917126c100aac159c2d7f0d4db5b6af3c9fa 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/tile-1.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/tile-1.c
@@ -1,5 +1,5 @@
 /* AMD GCN does not use 32-lane vectors, so the expected use counts mismatch.
-   { dg-skip-if "unsuitable dimensions" { openacc_amdgcn_accel_selected } { "*" } { "" } } */
+   { dg-skip-if "unsuitable dimensions" { openacc_radeon_accel_selected } { "*" } { "" } } */
 
 /* { dg-additional-options "-fopenacc-dim=32" } */
 

diff --git a/libgomp/testsuite/libgomp.oacc-c/c.exp b/libgomp/testsuite/libgomp.oacc-c/c.exp
index 7f13242fd59d6119167436b311a92de8f1ada394..48cbc980731b8bfbcea83e4a8949b159e71d0315 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-c/c.exp
+++ b/libgomp/testsuite/libgomp.oacc-c/c.exp
@@ -51,15 +51,6 @@ foreach offload_target [concat [split $offload_targets ","] "disable"] {
            unsupported "$subdir $offload_target offloading"
            continue
        }
-       gcn {
-           if { ![check_effective_target_openacc_amdgcn_accel_present] } {
-               # Don't bother; execution testing is going to FAIL.
-               untested "$subdir $offload_target offloading: supported, but hardware not accessible"
-               continue
-           }
-
-           set acc_mem_shared 0
-       }
        host {
            set acc_mem_shared 1
        }
@@ -78,6 +69,15 @@ foreach offload_target [concat [split $offload_targets ","] "disable"] {
 
            set acc_mem_shared 0
        }
+       radeon {
+           if { ![check_effective_target_openacc_radeon_accel_present] } {
+               # Don't bother; execution testing is going to FAIL.
+               untested "$subdir $offload_target offloading: supported, but hardware not accessible"
+               continue
+           }
+
+           set acc_mem_shared 0
+       }
        default {
            error "Unknown OpenACC device type: $openacc_device_type (offload target: $offload_target)"
        }

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/error_stop-1.f b/libgomp/testsuite/libgomp.oacc-fortran/error_stop-1.f
index e7358f4f20df08b521630e06ea8a7ee0d7f6e622..a3f54d57bc3562263285954b0476500387e82251 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/error_stop-1.f
+++ b/libgomp/testsuite/libgomp.oacc-fortran/error_stop-1.f
@@ -17,7 +17,7 @@
 ! In gfortran's main program, libfortran's set_options is called - which sets
 ! compiler_options.backtrace = 1 by default.  For an offload libgfortran, this
 ! is never called and, hence, "Error termination." is never printed.  Thus:
-! { dg-output "Error termination.*" { target { ! { openacc_nvidia_accel_selected || openacc_amdgcn_accel_selected } } } }
+! { dg-output "Error termination.*" { target { ! { openacc_nvidia_accel_selected || openacc_radeon_accel_selected } } } }
 !
 ! PR85463:
 ! { dg-output "libgomp: cuStreamSynchronize error.*" { target openacc_nvidia_accel_selected } }

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/error_stop-2.f b/libgomp/testsuite/libgomp.oacc-fortran/error_stop-2.f
index fca1d960f66d88f378ec6611338347b5dbc74a40..5d5d20d1bc53df87732e2ae003058cf825da32a7 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/error_stop-2.f
+++ b/libgomp/testsuite/libgomp.oacc-fortran/error_stop-2.f
@@ -17,7 +17,7 @@
 ! In gfortran's main program, libfortran's set_options is called - which sets
 ! compiler_options.backtrace = 1 by default.  For an offload libgfortran, this
 ! is never called and, hence, "Error termination." is never printed.  Thus:
-! { dg-output "Error termination.*" { target { ! { openacc_nvidia_accel_selected || openacc_amdgcn_accel_selected } } } }
+! { dg-output "Error termination.*" { target { ! { openacc_nvidia_accel_selected || openacc_radeon_accel_selected } } } }
 !
 ! PR85463:
 ! { dg-output "libgomp: cuStreamSynchronize error.*" { target openacc_nvidia_accel_selected } }

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/error_stop-3.f b/libgomp/testsuite/libgomp.oacc-fortran/error_stop-3.f
index 2ae0b0d160289b9b1e4a5ee612bade19f3e3b36b..edb063b182bc42127ef84a6357e374c1780bdd7a 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/error_stop-3.f
+++ b/libgomp/testsuite/libgomp.oacc-fortran/error_stop-3.f
@@ -17,7 +17,7 @@
 ! In gfortran's main program, libfortran's set_options is called - which sets
 ! compiler_options.backtrace = 1 by default.  For an offload libgfortran, this
 ! is never called and, hence, "Error termination." is never printed.  Thus:
-! { dg-output "Error termination.*" { target { ! { openacc_nvidia_accel_selected || openacc_amdgcn_accel_selected } } } }
+! { dg-output "Error termination.*" { target { ! { openacc_nvidia_accel_selected || openacc_radeon_accel_selected } } } }
 !
 ! PR85463:
 ! { dg-output "libgomp: cuStreamSynchronize error.*" { target openacc_nvidia_accel_selected } }

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/fortran.exp b/libgomp/testsuite/libgomp.oacc-fortran/fortran.exp
index 60f0889a07cffdd1ce4d380c56d494afae7387fa..d60790325053f60277cca560dec9653c8a109c30 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/fortran.exp
+++ b/libgomp/testsuite/libgomp.oacc-fortran/fortran.exp
@@ -82,8 +82,11 @@ if { $lang_test_file_found } {
                unsupported "$subdir $offload_target offloading"
                continue
            }
-           gcn {
-               if { ![check_effective_target_openacc_amdgcn_accel_present] } {
+           host {
+               set acc_mem_shared 1
+           }
+           nvidia {
+               if { ![check_effective_target_openacc_nvidia_accel_present] } {
                    # Don't bother; execution testing is going to FAIL.
                    untested "$subdir $offload_target offloading: supported, but hardware not accessible"
                    continue
@@ -91,11 +94,8 @@ if { $lang_test_file_found } {
 
                set acc_mem_shared 0
            }
-           host {
-               set acc_mem_shared 1
-           }
-           nvidia {
-               if { ![check_effective_target_openacc_nvidia_accel_present] } {
+           radeon {
+               if { ![check_effective_target_openacc_radeon_accel_present] } {
                    # Don't bother; execution testing is going to FAIL.
                    untested "$subdir $offload_target offloading: supported, but hardware not accessible"
                    continue