diff --git a/libgomp/ChangeLog b/libgomp/ChangeLog index 53a5511b033..41ee3ed60c5 100644 --- a/libgomp/ChangeLog +++ b/libgomp/ChangeLog @@ -1,5 +1,15 @@ 2019-02-22 Thomas Schwinge + * testsuite/libgomp.oacc-c++/c++.exp: Specify + "-foffload=$offload_target". + * testsuite/libgomp.oacc-c/c.exp: Likewise. + * testsuite/libgomp.oacc-fortran/fortran.exp: Likewise. + * testsuite/lib/libgomp.exp + (check_effective_target_openacc_nvidia_accel_configured): Remove, + as (conceptually) merged into + check_effective_target_openacc_nvidia_accel_selected. Adjust all + users. + * plugin/configfrag.ac: Populate and AC_SUBST offload_targets. * testsuite/libgomp-test-support.exp.in: Adjust. * testsuite/lib/libgomp.exp: Likewise. Don't populate diff --git a/libgomp/testsuite/lib/libgomp.exp b/libgomp/testsuite/lib/libgomp.exp index cb75e06c585..14d9b5f1305 100644 --- a/libgomp/testsuite/lib/libgomp.exp +++ b/libgomp/testsuite/lib/libgomp.exp @@ -362,18 +362,6 @@ proc check_effective_target_offload_device_shared_as { } { } ] } -# Return 1 if configured for 'nvptx' offloading. - -proc check_effective_target_openacc_nvidia_accel_configured { } { - global offload_targets - if { ![string match "*,nvptx*,*" ",$offload_targets,"] } { - return 0 - } - # PR libgomp/65099: Currently, we only support offloading in 64-bit - # configurations. - return [is-effective-target lp64] -} - # Return 1 if at least one Nvidia GPU is accessible. proc check_effective_target_openacc_nvidia_accel_present { } { diff --git a/libgomp/testsuite/libgomp.oacc-c++/c++.exp b/libgomp/testsuite/libgomp.oacc-c++/c++.exp index 29805c67bad..dcefa792ca4 100644 --- a/libgomp/testsuite/libgomp.oacc-c++/c++.exp +++ b/libgomp/testsuite/libgomp.oacc-c++/c++.exp @@ -112,6 +112,11 @@ if { $lang_test_file_found } { } set tagopt "$tagopt -DACC_MEM_SHARED=$acc_mem_shared" + # To avoid compilation overhead, and to keep simple '-foffload=[...]' + # handling in test cases, by default only build for the offload target + # that we're actually going to test. + set tagopt "$tagopt -foffload=$offload_target" + # Force usage of the corresponding OpenACC device type. setenv ACC_DEVICE_TYPE $openacc_device_type # To get better test coverage for device-specific code that is only 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 c94f268462f..fdf4eb08f8a 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_configured } } */ +/* { dg-excess-errors "lto1, mkoffload and lto-wrapper fatal errors" { target openacc_nvidia_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_configured } } */ +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 } } */ { var++; } diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-dims.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-dims.c index d7cd0461b53..7e699f476b2 100644 --- a/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-dims.c +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-dims.c @@ -154,7 +154,7 @@ int main () int gangs_min, gangs_max, workers_min, workers_max, vectors_min, vectors_max; gangs_min = workers_min = vectors_min = INT_MAX; gangs_max = workers_max = vectors_max = INT_MIN; -#pragma acc parallel copy (vectors_actual) /* { dg-warning "using vector_length \\(32\\), ignoring 1" "" { target openacc_nvidia_accel_configured } } */ \ +#pragma acc parallel copy (vectors_actual) /* { dg-warning "using vector_length \\(32\\), ignoring 1" "" { target openacc_nvidia_accel_selected } } */ \ vector_length (VECTORS) /* { dg-warning "'vector_length' value must be positive" "" { target c++ } } */ { /* We're actually executing with vector_length (1), just the GCC nvptx @@ -265,7 +265,7 @@ int main () int gangs_min, gangs_max, workers_min, workers_max, vectors_min, vectors_max; gangs_min = workers_min = vectors_min = INT_MAX; gangs_max = workers_max = vectors_max = INT_MIN; -#pragma acc parallel copy (workers_actual) /* { dg-warning "using num_workers \\(32\\), ignoring 2097152" "" { target openacc_nvidia_accel_configured } } */ \ +#pragma acc parallel copy (workers_actual) /* { dg-warning "using num_workers \\(32\\), ignoring 2097152" "" { target openacc_nvidia_accel_selected } } */ \ num_workers (WORKERS) { if (acc_on_device (acc_device_host)) @@ -350,7 +350,7 @@ int main () int gangs_min, gangs_max, workers_min, workers_max, vectors_min, vectors_max; gangs_min = workers_min = vectors_min = INT_MAX; gangs_max = workers_max = vectors_max = INT_MIN; -#pragma acc parallel copy (vectors_actual) /* { dg-warning "using vector_length \\(1024\\), ignoring 2097152" "" { target openacc_nvidia_accel_configured } } */ \ +#pragma acc parallel copy (vectors_actual) /* { dg-warning "using vector_length \\(1024\\), ignoring 2097152" "" { target openacc_nvidia_accel_selected } } */ \ vector_length (VECTORS) { if (acc_on_device (acc_device_host)) @@ -390,7 +390,7 @@ int main () int gangs_min, gangs_max, workers_min, workers_max, vectors_min, vectors_max; gangs_min = workers_min = vectors_min = INT_MAX; gangs_max = workers_max = vectors_max = INT_MIN; -#pragma acc parallel copy (vectors_actual) /* { dg-warning "using vector_length \\(32\\), ignoring runtime setting" "" { target openacc_nvidia_accel_configured } } */ \ +#pragma acc parallel copy (vectors_actual) /* { dg-warning "using vector_length \\(32\\), ignoring runtime setting" "" { target openacc_nvidia_accel_selected } } */ \ vector_length (vectors) { if (acc_on_device (acc_device_host)) @@ -437,7 +437,7 @@ int main () int gangs_min, gangs_max, workers_min, workers_max, vectors_min, vectors_max; gangs_min = workers_min = vectors_min = INT_MAX; gangs_max = workers_max = vectors_max = INT_MIN; -#pragma acc parallel copy (gangs_actual, workers_actual, vectors_actual) /* { dg-warning "using vector_length \\(32\\), ignoring 11" "" { target openacc_nvidia_accel_configured } } */ \ +#pragma acc parallel copy (gangs_actual, workers_actual, vectors_actual) /* { dg-warning "using vector_length \\(32\\), ignoring 11" "" { target openacc_nvidia_accel_selected } } */ \ num_gangs (gangs) \ num_workers (WORKERS) \ vector_length (VECTORS) diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/vector-length-64-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/vector-length-64-1.c index b6ee732f863..2d57ad4464a 100644 --- a/libgomp/testsuite/libgomp.oacc-c-c++-common/vector-length-64-1.c +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/vector-length-64-1.c @@ -4,7 +4,7 @@ int main (void) { -#pragma acc parallel vector_length (64) num_workers (16) /* { dg-warning "using num_workers \\(15\\), ignoring 16" "" { target openacc_nvidia_accel_configured } } */ +#pragma acc parallel vector_length (64) num_workers (16) /* { dg-warning "using num_workers \\(15\\), ignoring 16" "" { target openacc_nvidia_accel_selected } } */ { #pragma acc loop worker for (unsigned int i = 0; i < 32; i++) diff --git a/libgomp/testsuite/libgomp.oacc-c/c.exp b/libgomp/testsuite/libgomp.oacc-c/c.exp index 2cd75be0cc7..55cd40f1e99 100644 --- a/libgomp/testsuite/libgomp.oacc-c/c.exp +++ b/libgomp/testsuite/libgomp.oacc-c/c.exp @@ -75,6 +75,11 @@ foreach offload_target [concat [split $offload_targets ","] "disable"] { } set tagopt "$tagopt -DACC_MEM_SHARED=$acc_mem_shared" + # To avoid compilation overhead, and to keep simple '-foffload=[...]' + # handling in test cases, by default only build for the offload target + # that we're actually going to test. + set tagopt "$tagopt -foffload=$offload_target" + # Force usage of the corresponding OpenACC device type. setenv ACC_DEVICE_TYPE $openacc_device_type # To get better test coverage for device-specific code that is only diff --git a/libgomp/testsuite/libgomp.oacc-fortran/fortran.exp b/libgomp/testsuite/libgomp.oacc-fortran/fortran.exp index b2baa73d91a..af25a22a522 100644 --- a/libgomp/testsuite/libgomp.oacc-fortran/fortran.exp +++ b/libgomp/testsuite/libgomp.oacc-fortran/fortran.exp @@ -94,6 +94,11 @@ if { $lang_test_file_found } { } set tagopt "$tagopt -DACC_MEM_SHARED=$acc_mem_shared" + # To avoid compilation overhead, and to keep simple '-foffload=[...]' + # handling in test cases, by default only build for the offload target + # that we're actually going to test. + set tagopt "$tagopt -foffload=$offload_target" + # Force usage of the corresponding OpenACC device type. setenv ACC_DEVICE_TYPE $openacc_device_type # For Fortran we're doing torture testing, as Fortran has far more tests