From patchwork Tue Apr 21 12:24:46 2020 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 8bit X-Patchwork-Submitter: Thomas Schwinge X-Patchwork-Id: 1274203 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@bilbo.ozlabs.org Authentication-Results: ozlabs.org; spf=pass (sender SPF authorized) smtp.mailfrom=gcc.gnu.org (client-ip=8.43.85.97; helo=sourceware.org; envelope-from=gcc-patches-bounces@gcc.gnu.org; receiver=) Authentication-Results: ozlabs.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Received: from sourceware.org (server2.sourceware.org [8.43.85.97]) (using TLSv1.3 with cipher TLS_AES_256_GCM_SHA384 (256/256 bits) key-exchange X25519 server-signature RSA-PSS (4096 bits) server-digest SHA256) (No client certificate requested) by ozlabs.org (Postfix) with ESMTPS id 4962nt47SJz9sSX for ; Tue, 21 Apr 2020 22:25:05 +1000 (AEST) Received: from server2.sourceware.org (localhost [IPv6:::1]) by sourceware.org (Postfix) with ESMTP id 12A58386F425; Tue, 21 Apr 2020 12:25:02 +0000 (GMT) X-Original-To: gcc-patches@gcc.gnu.org Delivered-To: gcc-patches@gcc.gnu.org Received: from esa1.mentor.iphmx.com (esa1.mentor.iphmx.com [68.232.129.153]) by sourceware.org (Postfix) with ESMTPS id 29A48384B0C1 for ; Tue, 21 Apr 2020 12:24:58 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.3.2 sourceware.org 29A48384B0C1 Authentication-Results: sourceware.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Authentication-Results: sourceware.org; spf=pass smtp.mailfrom=Thomas_Schwinge@mentor.com IronPort-SDR: UMAmiCTo0rHo74uZigB+bfhz9/MRSoeSVGrHf9wfoCZWawOkuZpEeEq1dC7l+mlQ9Rmribgojk FAC9PH6k5VaiEM1MfE2K+uWktwfEACtjprcZslgWs4Z1EaeXaUUs50kQ5jSCYKy2hz5zKxNkDH imRwacCnagFr5tH7IOeUetjTdjJhZlvJFY5je7AzKR+PVMXgGTNbGnTLUbeeuLlEJTqWAYuNGG AlbCpL1A6HzKw1HNSTNAn30QilhIZv370BN9DbpHERyUBpggDVSq98u9tWAFypIidC1Q2ydF9j QiU= X-IronPort-AV: E=Sophos;i="5.72,410,1580803200"; d="scan'208,223";a="50067828" Received: from orw-gwy-02-in.mentorg.com ([192.94.38.167]) by esa1.mentor.iphmx.com with ESMTP; 21 Apr 2020 04:24:56 -0800 IronPort-SDR: B4Pc55igAPAtuw0ppZZPiVSsMtmKCFHZCjdcaVbsKckBIFcymW2WCzTcCKAa0zo97nNRVUu2SC VqbMnNZx342kFTI++0SH/adqKO8WGbsjY8W91TQ7joZi52VBBxu3v7WMWKVoL2pSJm06aOwmXG CuYgbf/GJj/ExFizdRTc/Cx6iIkoparVCWgRvWxIXaaM6l/B9+o6fozmFV4a1bQy6cWi1mm+q8 7ArAUZR9N9jBQEh4Nt2qoTVE9yhH6oBdjJFEKpidu3Huz4qGjDT5QkLMPzMbRF0Z/55wVy9KFn tZ0= From: Thomas Schwinge To: Andrew Stubbs Subject: [AMD GCN] Use 'radeon' for the environment variable 'ACC_DEVICE_TYPE' (was: [committed, amdgcn/openacc] Rename acc_device_gcn to acc_device_radeon) In-Reply-To: <1ea0084f-7ae2-c4c0-e1c1-92d36db5f23d@codesourcery.com> References: <87blsqlre6.fsf@euler.schwinge.homeip.net> <1ea0084f-7ae2-c4c0-e1c1-92d36db5f23d@codesourcery.com> User-Agent: Notmuch/0.29.1+93~g67ed7df (https://notmuchmail.org) Emacs/25.2.2 (x86_64-pc-linux-gnu) Date: Tue, 21 Apr 2020 14:24:46 +0200 Message-ID: MIME-Version: 1.0 X-Originating-IP: [137.202.0.90] X-ClientProxiedBy: SVR-IES-MBX-04.mgc.mentorg.com (139.181.222.4) To svr-ies-mbx-01.mgc.mentorg.com (139.181.222.1) X-Spam-Status: No, score=-25.8 required=5.0 tests=BAYES_00, GIT_PATCH_0, GIT_PATCH_1, GIT_PATCH_2, GIT_PATCH_3, HEADER_FROM_DIFFERENT_DOMAINS, KAM_DMARC_STATUS, KAM_SHORT, SPF_HELO_NONE, SPF_PASS, TXREP autolearn=ham autolearn_force=no version=3.4.2 X-Spam-Checker-Version: SpamAssassin 3.4.2 (2018-09-13) on server2.sourceware.org X-BeenThere: gcc-patches@gcc.gnu.org X-Mailman-Version: 2.1.29 Precedence: list List-Id: Gcc-patches mailing list List-Unsubscribe: , List-Archive: List-Post: List-Help: List-Subscribe: , Cc: gcc-patches@gcc.gnu.org Errors-To: gcc-patches-bounces@gcc.gnu.org Sender: "Gcc-patches" Hi Andrew! On 2020-01-17T18:18:41+0000, Andrew Stubbs wrote: > I've committed this patch, pre-approved by Thomas. It's basically just a > search and replace. > > I had originally invented acc_device_gcn for the device number, but as > Thomas pointed out the OpenACC documentation already uses > acc_device_radeon for AMD devices. Presumably this is what any existing > code will use, if anything, so we ought to be compatible. > > There's no official release using the "wrong" name, so I don't believe > we need to retain that name for any reason. ACK. > Rename acc_device_gcn to acc_device_radeon I wondered whether for symmetry, the GCC-internal 'GOMP_DEVICE_GCN', 'OFFLOAD_TARGET_TYPE_GCN' should also be renamed to '*_RADEON'? Or, going by example of '*_NVIDIA_PTX', name them '*_AMD_GCN'. Or, in fact then leave them as '*_GCN', given Julian's point in another thread that "CPU architectures or core implementations [sometimes shift] company allegiance". Thank you for listening to me thinking aloud. ;-) But more importantly, what about the user-visible 'gcn' in the 'ACC_DEVICE_TYPE' environment variable? As I'd quoted in : | Per OpenACC 3.0, A.1.2. "AMD | GPU Targets", for example, there is 'acc_device_radeon' ... which you've addressed... | (and "the | case-insensitive name 'radeon' for the environment variable | 'ACC_DEVICE_TYPE'"). ..., but this not yet? Please see the attached "[AMD GCN] Use 'radeon' for the environment variable 'ACC_DEVICE_TYPE'", completely untested. Will you please test and review that? If approving this patch, please respond with "Reviewed-by: NAME " so that your effort will be recorded in the commit log, see . Grüße Thomas ----------------- Mentor Graphics (Deutschland) GmbH, Arnulfstraße 201, 80634 München / Germany Registergericht München HRB 106955, Geschäftsführer: Thomas Heurung, Alexander Walter From d82df713bbb3401cc346bf07d61ad597d302d5a6 Mon Sep 17 00:00:00 2001 From: Thomas Schwinge Date: Tue, 21 Apr 2020 14:16:24 +0200 Subject: [PATCH] [AMD 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 | 4 +++- libgomp/testsuite/lib/libgomp.exp | 16 ++++++++-------- libgomp/testsuite/libgomp.oacc-c++/c++.exp | 18 +++++++++--------- .../libgomp.oacc-c++/firstprivate-mappings-1.C | 2 +- .../acc_get_property-gcn.c | 2 +- .../asyncwait-nop-1.c | 2 +- .../firstprivate-mappings-1.c | 2 +- .../function-not-offloaded.c | 4 ++-- .../libgomp.oacc-c-c++-common/loop-auto-1.c | 2 +- .../loop-dim-default.c | 2 +- .../libgomp.oacc-c-c++-common/routine-wv-2.c | 2 +- .../libgomp.oacc-c-c++-common/tile-1.c | 2 +- libgomp/testsuite/libgomp.oacc-c/c.exp | 18 +++++++++--------- .../libgomp.oacc-fortran/error_stop-1.f | 2 +- .../libgomp.oacc-fortran/error_stop-2.f | 2 +- .../libgomp.oacc-fortran/error_stop-3.f | 2 +- .../testsuite/libgomp.oacc-fortran/fortran.exp | 14 +++++++------- 17 files changed, 49 insertions(+), 47 deletions(-) diff --git a/libgomp/oacc-init.c b/libgomp/oacc-init.c index ef12b4c16d0..5d786a5a2e7 100644 --- 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 e7ce784314d..ee5f0e57b19 100644 --- 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 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 c06c2a097e3..7200ec19c47 100644 --- 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 7b3e670073c..b046bf2912d 100644 --- 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 ce59264a60d..4b1fb5e0e76 100644 --- 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 #include 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 754e015a280..7496426c8fa 100644 --- 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 253f8bf0bd0..2cdd2d1525e 100644 --- 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 517004a562d..64f8ab812a6 100644 --- 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 0c9ae957460..0273c2bddd7 100644 --- 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 30f0539707f..ca771646655 100644 --- 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 609f9f6a7da..9769ee72430 100644 --- 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 #include -#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 c019fe55c7a..5757917126c 100644 --- 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 7f13242fd59..48cbc980731 100644 --- 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 e7358f4f20d..a3f54d57bc3 100644 --- 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 fca1d960f66..5d5d20d1bc5 100644 --- 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 2ae0b0d1602..edb063b182b 100644 --- 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 60f0889a07c..d6079032505 100644 --- 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 -- 2.17.1