Hi! On 2019-12-10T15:23:01+0100, Frederik Harwath <frede...@codesourcery.com> wrote: > On 09.12.19 16:58, Harwath, Frederik wrote: >> [use] the location of clauses in warnings instead of the location of the >> loop to which the clause belongs.
> Frederik Harwath (2): > Use clause locations in OpenACC nested reduction warnings Basically: -warning_at (gimple_location (stmt), 0, +warning_at (OMP_CLAUSE_LOCATION (clause), 0, > Add tests to verify OpenACC clause locations Similar changes are desirable for other directives/clauses, too. I've just pushed "[OpenACC] More precise diagnostics for 'gang', 'worker', 'vector' clauses with arguments on 'loop' only allowed in 'kernels' regions" to master branch in commit beddd1762ad2bbe84dd776c54489153f83f21e56, and backported to releases/gcc-10 in commit 8d09f49006ce4c2f8d4018206c12e131c49ca6ce, see attached. 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 beddd1762ad2bbe84dd776c54489153f83f21e56 Mon Sep 17 00:00:00 2001 From: Thomas Schwinge <tho...@codesourcery.com> Date: Tue, 27 Oct 2020 17:13:16 +0100 Subject: [PATCH] [OpenACC] More precise diagnostics for 'gang', 'worker', 'vector' clauses with arguments on 'loop' only allowed in 'kernels' regions Instead of at the location of the 'loop' directive, 'error_at' the location of the improper clause, and 'inform' at the location of the enclosing parent compute construct/routine. The Fortran testcases come with some XFAILing, to be resolved later. gcc/ * omp-low.c (scan_omp_for) <OpenACC>: More precise diagnostics for 'gang', 'worker', 'vector' clauses with arguments only allowed in 'kernels' regions. gcc/testsuite/ * c-c++-common/goacc/pr92793-1.c: Extend. * gfortran.dg/goacc/pr92793-1.f90: Likewise. --- gcc/omp-low.c | 29 +++++++---- gcc/testsuite/c-c++-common/goacc/pr92793-1.c | 37 +++++++++++++ gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 | 52 +++++++++++++++++++ 3 files changed, 108 insertions(+), 10 deletions(-) diff --git a/gcc/omp-low.c b/gcc/omp-low.c index 5392fa7e3086..de5142f979b0 100644 --- a/gcc/omp-low.c +++ b/gcc/omp-low.c @@ -2418,30 +2418,39 @@ scan_omp_for (gomp_for *stmt, omp_context *outer_ctx) if (!tgt || is_oacc_parallel_or_serial (tgt)) for (tree c = clauses; c; c = OMP_CLAUSE_CHAIN (c)) { - char const *check = NULL; - + tree c_op0; switch (OMP_CLAUSE_CODE (c)) { case OMP_CLAUSE_GANG: - check = "gang"; + c_op0 = OMP_CLAUSE_GANG_EXPR (c); break; case OMP_CLAUSE_WORKER: - check = "worker"; + c_op0 = OMP_CLAUSE_WORKER_EXPR (c); break; case OMP_CLAUSE_VECTOR: - check = "vector"; + c_op0 = OMP_CLAUSE_VECTOR_EXPR (c); break; default: - break; + continue; } - if (check && OMP_CLAUSE_OPERAND (c, 0)) - error_at (gimple_location (stmt), - "argument not permitted on %qs clause in" - " OpenACC %<parallel%> or %<serial%>", check); + if (c_op0) + { + error_at (OMP_CLAUSE_LOCATION (c), + "argument not permitted on %qs clause", + omp_clause_code_name[OMP_CLAUSE_CODE (c)]); + if (tgt) + inform (gimple_location (outer_ctx->stmt), + "enclosing parent compute construct"); + else if (oacc_get_fn_attrib (current_function_decl)) + inform (DECL_SOURCE_LOCATION (current_function_decl), + "enclosing routine"); + else + gcc_unreachable (); + } } if (tgt && is_oacc_kernels (tgt)) diff --git a/gcc/testsuite/c-c++-common/goacc/pr92793-1.c b/gcc/testsuite/c-c++-common/goacc/pr92793-1.c index d7a2ae487992..77ebb20265cf 100644 --- a/gcc/testsuite/c-c++-common/goacc/pr92793-1.c +++ b/gcc/testsuite/c-c++-common/goacc/pr92793-1.c @@ -54,3 +54,40 @@ reduction(-:sum ) /* { dg-line sum2 } */ \ } } } + + +void +a_sl() { +#pragma acc serial loop /* { dg-message "9: enclosing parent compute construct" } */ \ + gang(num:5) /* { dg-error "5: argument not permitted on 'gang' clause" } */ \ + worker(num:5) /* { dg-error "3: argument not permitted on 'worker' clause" } */ \ + vector(length:5) /* { dg-error "4: argument not permitted on 'vector' clause" } */ + for (int i = 0; i < 10; i++) + ; +} + +void +a_s_l() { +#pragma acc serial /* { dg-message "9: enclosing parent compute construct" } */ + { +#pragma acc loop \ + gang(num:5) /* { dg-error "8: argument not permitted on 'gang' clause" } */ \ + worker(num:5) /* { dg-error "4: argument not permitted on 'worker' clause" } */ \ + vector(length:5) /* { dg-error "3: argument not permitted on 'vector' clause" } */ + for (int i = 0; i < 10; i++) + ; + } +} + +void a_r(); +#pragma acc routine(a_r) + +void +a_r() { /* { dg-message "1: enclosing routine" } */ +#pragma acc loop \ + gang(num:5) /* { dg-error "4: argument not permitted on 'gang' clause" } */ \ + worker(num:5) /* { dg-error "5: argument not permitted on 'worker' clause" } */ \ + vector(length:5) /* { dg-error "3: argument not permitted on 'vector' clause" } */ + for (int i = 0; i < 10; i++) + ; +} diff --git a/gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 b/gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 index 72dd6b7b8f81..23d6886580a1 100644 --- a/gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 +++ b/gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 @@ -45,3 +45,55 @@ subroutine check () end do !$acc end parallel end subroutine check + + +subroutine gwv_sl () + implicit none (type, external) + integer :: i + + !$acc serial loop & + !$acc & gang(num:5) & ! { dg-error "argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } .+6 } + !$acc & worker(num:5) & ! { dg-error "argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } .+4 } + !$acc & vector(length:5) & ! { dg-error "argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } } + !$acc & ! { dg-bogus "14: argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } .+2 } + ! { dg-message "99: enclosing parent compute construct" "" { target *-*-* } .-1 } + do i = 0, 10 + end do + !$acc end serial loop +end subroutine gwv_sl + +subroutine gwv_s_l () + implicit none (type, external) + integer :: i + + !$acc serial ! { dg-message "72: enclosing parent compute construct" } + !$acc loop & + !$acc & gang(num:5) & ! { dg-error "argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } .+5 } + !$acc & worker(num:5) & ! { dg-error "argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } .+3 } + !$acc & vector(length:5) & ! { dg-error "argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } } + !$acc & ! { dg-bogus "14: argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } .+1 } + do i = 0, 10 + end do + !$acc end serial +end subroutine gwv_s_l + +subroutine gwv_r () ! { dg-message "16: enclosing routine" } + implicit none (type, external) + integer :: i + + !$acc routine(gwv_r) + + !$acc loop & + !$acc & gang(num:5) & ! { dg-error "argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } .+5 } + !$acc & worker(num:5) & ! { dg-error "argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } .+3 } + !$acc & vector(length:5) & ! { dg-error "argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } } + !$acc & ! { dg-bogus "14: argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } .+1 } + do i = 0, 10 + end do +end subroutine gwv_r -- 2.17.1
>From 8d09f49006ce4c2f8d4018206c12e131c49ca6ce Mon Sep 17 00:00:00 2001 From: Thomas Schwinge <tho...@codesourcery.com> Date: Tue, 27 Oct 2020 17:13:16 +0100 Subject: [PATCH] [OpenACC] More precise diagnostics for 'gang', 'worker', 'vector' clauses with arguments on 'loop' only allowed in 'kernels' regions Instead of at the location of the 'loop' directive, 'error_at' the location of the improper clause, and 'inform' at the location of the enclosing parent compute construct/routine. The Fortran testcases come with some XFAILing, to be resolved later. gcc/ * omp-low.c (scan_omp_for) <OpenACC>: More precise diagnostics for 'gang', 'worker', 'vector' clauses with arguments only allowed in 'kernels' regions. gcc/testsuite/ * c-c++-common/goacc/pr92793-1.c: Extend. * gfortran.dg/goacc/pr92793-1.f90: Likewise. (cherry picked from commit beddd1762ad2bbe84dd776c54489153f83f21e56) --- gcc/omp-low.c | 29 +++++++---- gcc/testsuite/c-c++-common/goacc/pr92793-1.c | 37 +++++++++++++ gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 | 52 +++++++++++++++++++ 3 files changed, 108 insertions(+), 10 deletions(-) diff --git a/gcc/omp-low.c b/gcc/omp-low.c index 94c7ef33757e..16afc3720af2 100644 --- a/gcc/omp-low.c +++ b/gcc/omp-low.c @@ -2411,30 +2411,39 @@ scan_omp_for (gomp_for *stmt, omp_context *outer_ctx) if (!tgt || is_oacc_parallel_or_serial (tgt)) for (tree c = clauses; c; c = OMP_CLAUSE_CHAIN (c)) { - char const *check = NULL; - + tree c_op0; switch (OMP_CLAUSE_CODE (c)) { case OMP_CLAUSE_GANG: - check = "gang"; + c_op0 = OMP_CLAUSE_GANG_EXPR (c); break; case OMP_CLAUSE_WORKER: - check = "worker"; + c_op0 = OMP_CLAUSE_WORKER_EXPR (c); break; case OMP_CLAUSE_VECTOR: - check = "vector"; + c_op0 = OMP_CLAUSE_VECTOR_EXPR (c); break; default: - break; + continue; } - if (check && OMP_CLAUSE_OPERAND (c, 0)) - error_at (gimple_location (stmt), - "argument not permitted on %qs clause in" - " OpenACC %<parallel%> or %<serial%>", check); + if (c_op0) + { + error_at (OMP_CLAUSE_LOCATION (c), + "argument not permitted on %qs clause", + omp_clause_code_name[OMP_CLAUSE_CODE (c)]); + if (tgt) + inform (gimple_location (outer_ctx->stmt), + "enclosing parent compute construct"); + else if (oacc_get_fn_attrib (current_function_decl)) + inform (DECL_SOURCE_LOCATION (current_function_decl), + "enclosing routine"); + else + gcc_unreachable (); + } } if (tgt && is_oacc_kernels (tgt)) diff --git a/gcc/testsuite/c-c++-common/goacc/pr92793-1.c b/gcc/testsuite/c-c++-common/goacc/pr92793-1.c index d7a2ae487992..77ebb20265cf 100644 --- a/gcc/testsuite/c-c++-common/goacc/pr92793-1.c +++ b/gcc/testsuite/c-c++-common/goacc/pr92793-1.c @@ -54,3 +54,40 @@ reduction(-:sum ) /* { dg-line sum2 } */ \ } } } + + +void +a_sl() { +#pragma acc serial loop /* { dg-message "9: enclosing parent compute construct" } */ \ + gang(num:5) /* { dg-error "5: argument not permitted on 'gang' clause" } */ \ + worker(num:5) /* { dg-error "3: argument not permitted on 'worker' clause" } */ \ + vector(length:5) /* { dg-error "4: argument not permitted on 'vector' clause" } */ + for (int i = 0; i < 10; i++) + ; +} + +void +a_s_l() { +#pragma acc serial /* { dg-message "9: enclosing parent compute construct" } */ + { +#pragma acc loop \ + gang(num:5) /* { dg-error "8: argument not permitted on 'gang' clause" } */ \ + worker(num:5) /* { dg-error "4: argument not permitted on 'worker' clause" } */ \ + vector(length:5) /* { dg-error "3: argument not permitted on 'vector' clause" } */ + for (int i = 0; i < 10; i++) + ; + } +} + +void a_r(); +#pragma acc routine(a_r) + +void +a_r() { /* { dg-message "1: enclosing routine" } */ +#pragma acc loop \ + gang(num:5) /* { dg-error "4: argument not permitted on 'gang' clause" } */ \ + worker(num:5) /* { dg-error "5: argument not permitted on 'worker' clause" } */ \ + vector(length:5) /* { dg-error "3: argument not permitted on 'vector' clause" } */ + for (int i = 0; i < 10; i++) + ; +} diff --git a/gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 b/gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 index 72dd6b7b8f81..23d6886580a1 100644 --- a/gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 +++ b/gcc/testsuite/gfortran.dg/goacc/pr92793-1.f90 @@ -45,3 +45,55 @@ subroutine check () end do !$acc end parallel end subroutine check + + +subroutine gwv_sl () + implicit none (type, external) + integer :: i + + !$acc serial loop & + !$acc & gang(num:5) & ! { dg-error "argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } .+6 } + !$acc & worker(num:5) & ! { dg-error "argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } .+4 } + !$acc & vector(length:5) & ! { dg-error "argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } } + !$acc & ! { dg-bogus "14: argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } .+2 } + ! { dg-message "99: enclosing parent compute construct" "" { target *-*-* } .-1 } + do i = 0, 10 + end do + !$acc end serial loop +end subroutine gwv_sl + +subroutine gwv_s_l () + implicit none (type, external) + integer :: i + + !$acc serial ! { dg-message "72: enclosing parent compute construct" } + !$acc loop & + !$acc & gang(num:5) & ! { dg-error "argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } .+5 } + !$acc & worker(num:5) & ! { dg-error "argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } .+3 } + !$acc & vector(length:5) & ! { dg-error "argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } } + !$acc & ! { dg-bogus "14: argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } .+1 } + do i = 0, 10 + end do + !$acc end serial +end subroutine gwv_s_l + +subroutine gwv_r () ! { dg-message "16: enclosing routine" } + implicit none (type, external) + integer :: i + + !$acc routine(gwv_r) + + !$acc loop & + !$acc & gang(num:5) & ! { dg-error "argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'gang' clause" "TODO" { xfail *-*-* } .+5 } + !$acc & worker(num:5) & ! { dg-error "argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } } + !$acc & & ! { dg-bogus "14: argument not permitted on 'worker' clause" "TODO" { xfail *-*-* } .+3 } + !$acc & vector(length:5) & ! { dg-error "argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } } + !$acc & ! { dg-bogus "14: argument not permitted on 'vector' clause" "TODO" { xfail *-*-* } .+1 } + do i = 0, 10 + end do +end subroutine gwv_r -- 2.17.1