[PING,PR,c/71381] C/C++ OpenACC cache directive rejects valid syntax

Message ID	87inxhnxm3.fsf@kepler.schwinge.homeip.net
State	New
Headers	show Return-Path: <gcc-patches-return-429516-incoming=patchwork.ozlabs.org@gcc.gnu.org> DomainKey-Signature: a=rsa-sha1; c=nofws; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender:from :to:cc:subject:in-reply-to:references:date:message-id :mime-version:content-type; q=dns; s=default; b=P5vJxbWBJstSOTya ufMPGDVMSZXf/JI2uYlYFqxBxFjG4vUf+3GiAOrQ8/LxksKJmaGvQH7hzZfnpazV 7gn1H+HeFBbHvtBZNG5fgGM6KAjg646FNLKx8iSgdkuUIR3cpobI6WCU3JgfUs8H Veiij3A50+DjG71yLugEzzOwxNc= Mailing-List: contact gcc-patches-help@gcc.gnu.org; run by ezmlm Precedence: bulk Sender: gcc-patches-owner@gcc.gnu.org From: Thomas Schwinge <thomas@codesourcery.com> To: <gcc-patches@gcc.gnu.org>, Jakub Jelinek <jakub@redhat.com> CC: James Norris <jnorris@codesourcery.com> Subject: Re: [PING] [PR c/71381] C/C++ OpenACC cache directive rejects valid syntax In-Reply-To: <20160608140717.GE7387@tucnak.redhat.com> References: <87d291ehls.fsf@kepler.schwinge.homeip.net> <877fe7sthf.fsf@kepler.schwinge.homeip.net> <87wplzrequ.fsf@hertz.schwinge.homeip.net> <20160608140717.GE7387@tucnak.redhat.com> User-Agent: Notmuch/0.9-125-g4686d11 (http://notmuchmail.org) Emacs/24.5.1 (i586-pc-linux-gnu) Date: Fri, 10 Jun 2016 12:31:48 +0200 Message-ID: <87inxhnxm3.fsf@kepler.schwinge.homeip.net> MIME-Version: 1.0 Content-Type: multipart/signed; boundary="=-=-="; micalg=pgp-sha1; protocol="application/pgp-signature"

diff --git gcc/ChangeLog gcc/ChangeLog index 3f7240c..6afbae7 100644 --- gcc/ChangeLog +++ gcc/ChangeLog @@ -1,3 +1,8 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + * omp-low.c (scan_sharing_clauses): Don't expect + OMP_CLAUSE__CACHE_. + 2016-06-10 Alan Hayward <alan.hayward@arm.com> PR tree-optimization/71407 diff --git gcc/c/ChangeLog gcc/c/ChangeLog index 4d30245..cd9f230 100644 --- gcc/c/ChangeLog +++ gcc/c/ChangeLog @@ -1,3 +1,9 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + * c-parser.c (c_parser_omp_variable_list) <OMP_CLAUSE__CACHE_>: + Loosen checking. + 2016-06-08 Martin Sebor <msebor@redhat.com> Jakub Jelinek <jakub@redhat.com> diff --git gcc/c/c-parser.c gcc/c/c-parser.c index 2fef1ac..94078a9 100644 --- gcc/c/c-parser.c +++ gcc/c/c-parser.c @@ -10613,6 +10613,8 @@ c_parser_omp_variable_list (c_parser *parser, switch (kind) { case OMP_CLAUSE__CACHE_: + /* The OpenACC cache directive explicitly only allows "array + elements or subarrays". */ if (c_parser_peek_token (parser)->type != CPP_OPEN_SQUARE) { c_parser_error (parser, "expected %<[%>"); @@ -10678,25 +10680,6 @@ c_parser_omp_variable_list (c_parser *parser, break; } - if (kind == OMP_CLAUSE__CACHE_) - { - if (TREE_CODE (low_bound) != INTEGER_CST - && !TREE_READONLY (low_bound)) - { - error_at (clause_loc, - "%qD is not a constant", low_bound); - t = error_mark_node; - } - - if (TREE_CODE (length) != INTEGER_CST - && !TREE_READONLY (length)) - { - error_at (clause_loc, - "%qD is not a constant", length); - t = error_mark_node; - } - } - t = tree_cons (low_bound, length, t); } break; diff --git gcc/cp/ChangeLog gcc/cp/ChangeLog index a94cebe..53d55d3 100644 --- gcc/cp/ChangeLog +++ gcc/cp/ChangeLog @@ -1,3 +1,9 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + * parser.c (cp_parser_omp_var_list_no_open) <OMP_CLAUSE__CACHE_>: + Loosen checking. + 2016-06-09 Paolo Carlini <paolo.carlini@oracle.com> PR c++/71465 diff --git gcc/cp/parser.c gcc/cp/parser.c index e01353d..632b25f 100644 --- gcc/cp/parser.c +++ gcc/cp/parser.c @@ -29984,6 +29984,8 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, switch (kind) { case OMP_CLAUSE__CACHE_: + /* The OpenACC cache directive explicitly only allows "array + elements or subarrays". */ if (cp_lexer_peek_token (parser->lexer)->type != CPP_OPEN_SQUARE) { error_at (token->location, "expected %<[%>"); @@ -30035,25 +30037,6 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, RT_CLOSE_SQUARE)) goto skip_comma; - if (kind == OMP_CLAUSE__CACHE_) - { - if (TREE_CODE (low_bound) != INTEGER_CST - && !TREE_READONLY (low_bound)) - { - error_at (token->location, - "%qD is not a constant", low_bound); - decl = error_mark_node; - } - - if (TREE_CODE (length) != INTEGER_CST - && !TREE_READONLY (length)) - { - error_at (token->location, - "%qD is not a constant", length); - decl = error_mark_node; - } - } - decl = tree_cons (low_bound, length, decl); } break; diff --git gcc/fortran/ChangeLog gcc/fortran/ChangeLog index b2354e7..67bc9e8 100644 --- gcc/fortran/ChangeLog +++ gcc/fortran/ChangeLog @@ -1,3 +1,8 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + * openmp.c (gfc_match_oacc_cache): Add comment. + 2016-06-05 Jerry DeLisle <jvdelisle@gcc.gnu.org> PR fortran/71404 diff --git gcc/fortran/openmp.c gcc/fortran/openmp.c index 2689d30..2c92794 100644 --- gcc/fortran/openmp.c +++ gcc/fortran/openmp.c @@ -1688,6 +1688,10 @@ match gfc_match_oacc_cache (void) { gfc_omp_clauses *c = gfc_get_omp_clauses (); + /* The OpenACC cache directive explicitly only allows "array elements or + subarrays", which we're currently not checking here. Either check this + after the call of gfc_match_omp_variable_list, or add something like a + only_sections variant next to its allow_sections parameter. */ match m = gfc_match_omp_variable_list (" (", &c->lists[OMP_LIST_CACHE], true, NULL, NULL, true); diff --git gcc/omp-low.c gcc/omp-low.c index 77bdb18..91d5fcf 100644 --- gcc/omp-low.c +++ gcc/omp-low.c @@ -2201,9 +2201,6 @@ scan_sharing_clauses (tree clauses, omp_context *ctx, break; case OMP_CLAUSE__CACHE_: - sorry ("Clause not supported yet"); - break; - default: gcc_unreachable (); } @@ -2368,9 +2365,6 @@ scan_sharing_clauses (tree clauses, omp_context *ctx, break; case OMP_CLAUSE__CACHE_: - sorry ("Clause not supported yet"); - break; - default: gcc_unreachable (); } diff --git gcc/testsuite/ChangeLog gcc/testsuite/ChangeLog index ae3f462..e15b009 100644 --- gcc/testsuite/ChangeLog +++ gcc/testsuite/ChangeLog @@ -1,3 +1,16 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + * c-c++-common/goacc/cache-1.c: Update. Move invalid usage tests + to... + * c-c++-common/goacc/cache-2.c: ... this new file. + * gfortran.dg/goacc/cache-1.f95: Move invalid usage tests to... + * gfortran.dg/goacc/cache-2.f95: ... this new file. + * gfortran.dg/goacc/coarray.f95: Update OpenACC cache directive + usage. + * gfortran.dg/goacc/cray.f95: Likewise. + * gfortran.dg/goacc/loop-1.f95: Likewise. + 2016-06-10 Alan Hayward <alan.hayward@arm.com> PR tree-optimization/71407 diff --git gcc/testsuite/c-c++-common/goacc/cache-1.c gcc/testsuite/c-c++-common/goacc/cache-1.c index 9503341..1d4759e 100644 --- gcc/testsuite/c-c++-common/goacc/cache-1.c +++ gcc/testsuite/c-c++-common/goacc/cache-1.c @@ -1,3 +1,7 @@ +/* OpenACC cache directive: valid usage. */ +/* For execution testing, this file is "#include"d from + libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c. */ + int main (int argc, char **argv) { @@ -21,57 +25,31 @@ main (int argc, char **argv) int n = 1; const int len = n; -#pragma acc cache /* { dg-error "expected '\\$' before end of line" } */ - -#pragma acc cache a[0:N] /* { dg-error "expected '\\\(' before 'a'" } */ - /* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } 26 } */ - -#pragma acc cache (a) /* { dg-error "expected '\\\['" } */ - -#pragma acc cache ( /* { dg-error "expected (identifier|unqualified-id) before end of line" } */ - -#pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\$' token" } */ - -#pragma acc cache (,) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" } */ - -#pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - -#pragma acc cache (a[0:N],) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" "" { xfail c } } */ - -#pragma acc cache (a[0:N]) copyin (a[0:N]) /* { dg-error "expected end of line before 'copyin'" } */ - -#pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\\)' token" } */ - -#pragma acc cache (a[0:N] b[0:N]) /* { dg-error "expected '\\\)' before 'b'" } */ - -#pragma acc cache (a[0:N] b[0:N}) /* { dg-error "expected '\\\)' before 'b'" } */ - /* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } 47 } */ - -#pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - -#pragma acc cache (a[ii]) /* { dg-error "'ii' is not a constant" } */ - -#pragma acc cache (a[idx:n]) /* { dg-error "'n' is not a constant" } */ - -#pragma acc cache (a[0:N]) ( /* { dg-error "expected end of line before '\$' token" } */ - -#pragma acc cache (a[0:N]) ii /* { dg-error "expected end of line before 'ii'" } */ - -#pragma acc cache (a[0:N] ii) /* { dg-error "expected '\$' before 'ii'" } */ - + /* Have at it, GCC! */ #pragma acc cache (a[0:N]) - #pragma acc cache (a[0:N], a[0:N]) - #pragma acc cache (a[0:N], b[0:N]) - #pragma acc cache (a[0]) - #pragma acc cache (a[0], a[1], b[0:N]) - +#pragma acc cache (a[i - 5]) +#pragma acc cache (a[i + 5:len]) +#pragma acc cache (a[i + 5:len - 1]) +#pragma acc cache (b[i]) +#pragma acc cache (b[i:len]) +#pragma acc cache (a[ii]) +#pragma acc cache (a[ii:len]) +#pragma acc cache (b[ii - 1]) +#pragma acc cache (b[ii - 1:len]) +#pragma acc cache (b[i - ii + 1]) +#pragma acc cache (b[i + ii - 1:len]) +#pragma acc cache (b[i * ii - 1:len + 1]) +#pragma acc cache (a[idx + 2]) +#pragma acc cache (a[idx:len + 2]) #pragma acc cache (a[idx]) - #pragma acc cache (a[idx:len]) +#pragma acc cache (a[idx + 2:len]) +#pragma acc cache (a[idx + 2 + i:len]) +#pragma acc cache (a[idx + 2 + i + ii:len]) b[ii] = a[ii]; } diff --git gcc/testsuite/c-c++-common/goacc/cache-1.c gcc/testsuite/c-c++-common/goacc/cache-2.c similarity index 83% copy from gcc/testsuite/c-c++-common/goacc/cache-1.c copy to gcc/testsuite/c-c++-common/goacc/cache-2.c index 9503341..f717515 100644 --- gcc/testsuite/c-c++-common/goacc/cache-1.c +++ gcc/testsuite/c-c++-common/goacc/cache-2.c @@ -1,3 +1,5 @@ +/* OpenACC cache directive: invalid usage. */ + int main (int argc, char **argv) { @@ -22,57 +24,24 @@ main (int argc, char **argv) const int len = n; #pragma acc cache /* { dg-error "expected '\\$' before end of line" } */ - #pragma acc cache a[0:N] /* { dg-error "expected '\\\(' before 'a'" } */ - /* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } 26 } */ - + /* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } 27 } */ #pragma acc cache (a) /* { dg-error "expected '\\\['" } */ - #pragma acc cache ( /* { dg-error "expected (identifier|unqualified-id) before end of line" } */ - #pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\$' token" } */ - #pragma acc cache (,) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" } */ - #pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - #pragma acc cache (a[0:N],) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" "" { xfail c } } */ - #pragma acc cache (a[0:N]) copyin (a[0:N]) /* { dg-error "expected end of line before 'copyin'" } */ - #pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\\)' token" } */ - #pragma acc cache (a[0:N] b[0:N]) /* { dg-error "expected '\\\)' before 'b'" } */ - #pragma acc cache (a[0:N] b[0:N}) /* { dg-error "expected '\\\)' before 'b'" } */ - /* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } 47 } */ - + /* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } 38 } */ #pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - -#pragma acc cache (a[ii]) /* { dg-error "'ii' is not a constant" } */ - -#pragma acc cache (a[idx:n]) /* { dg-error "'n' is not a constant" } */ - #pragma acc cache (a[0:N]) ( /* { dg-error "expected end of line before '\$' token" } */ - #pragma acc cache (a[0:N]) ii /* { dg-error "expected end of line before 'ii'" } */ - #pragma acc cache (a[0:N] ii) /* { dg-error "expected '\$' before 'ii'" } */ -#pragma acc cache (a[0:N]) - -#pragma acc cache (a[0:N], a[0:N]) - -#pragma acc cache (a[0:N], b[0:N]) - -#pragma acc cache (a[0]) - -#pragma acc cache (a[0], a[1], b[0:N]) - -#pragma acc cache (a[idx]) - -#pragma acc cache (a[idx:len]) - b[ii] = a[ii]; } } diff --git gcc/testsuite/gfortran.dg/goacc/cache-1.f95 gcc/testsuite/gfortran.dg/goacc/cache-1.f95 index 2aa9e05..39fbf2c 100644 --- gcc/testsuite/gfortran.dg/goacc/cache-1.f95 +++ gcc/testsuite/gfortran.dg/goacc/cache-1.f95 @@ -1,4 +1,6 @@ -! { dg-do compile } +! OpenACC cache directive: valid usage. +! For execution testing, this file is "#include"d from +! libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95. ! { dg-additional-options "-std=f2008" } program test @@ -6,11 +8,8 @@ program test integer :: i, d(10), e(5,13) do concurrent (i=1:5) - !$acc cache (d) !$acc cache (d(1:3)) !$acc cache (d(i:i+2)) - - !$acc cache (e) !$acc cache (e(1:3,2:4)) !$acc cache (e(i:i+2,i+1:i+3)) enddo diff --git gcc/testsuite/gfortran.dg/goacc/cache-2.f95 gcc/testsuite/gfortran.dg/goacc/cache-2.f95 new file mode 100644 index 0000000..be81878 --- /dev/null +++ gcc/testsuite/gfortran.dg/goacc/cache-2.f95 @@ -0,0 +1,12 @@ +! OpenACC cache directive: invalid usage. +! { dg-additional-options "-std=f2008" } + +program test + implicit none + integer :: i, d(10), e(5,13) + + do concurrent (i=1:5) + !$acc cache (d) ! { dg-error "" "TODO" { xfail *-*-* } } + !$acc cache (e) ! { dg-error "" "TODO" { xfail *-*-* } } + enddo +end diff --git gcc/testsuite/gfortran.dg/goacc/coarray.f95 gcc/testsuite/gfortran.dg/goacc/coarray.f95 index 932e1f7..f30917b8 100644 --- gcc/testsuite/gfortran.dg/goacc/coarray.f95 +++ gcc/testsuite/gfortran.dg/goacc/coarray.f95 @@ -24,7 +24,7 @@ contains !$acc end parallel loop !$acc parallel loop do i = 1,5 - !$acc cache (a) + !$acc cache (a) ! { dg-error "" "TODO" { xfail *-*-* } } enddo !$acc end parallel loop !$acc update device (a) diff --git gcc/testsuite/gfortran.dg/goacc/cray.f95 gcc/testsuite/gfortran.dg/goacc/cray.f95 index a35ab0d..705c18c 100644 --- gcc/testsuite/gfortran.dg/goacc/cray.f95 +++ gcc/testsuite/gfortran.dg/goacc/cray.f95 @@ -44,7 +44,8 @@ contains !$acc end parallel loop !$acc parallel loop do i = 1,5 - !$acc cache (ptr) ! TODO: This must fail, as in openacc-1_0-branch + !TODO: This must fail, as in openacc-1_0-branch. + !$acc cache (ptr) ! { dg-error "" "TODO" { xfail *-*-* } } enddo !$acc end parallel loop !$acc update device (ptr) diff --git gcc/testsuite/gfortran.dg/goacc/loop-1.f95 gcc/testsuite/gfortran.dg/goacc/loop-1.f95 index b5f9e03..a605f03 100644 --- gcc/testsuite/gfortran.dg/goacc/loop-1.f95 +++ gcc/testsuite/gfortran.dg/goacc/loop-1.f95 @@ -158,15 +158,16 @@ subroutine test1 enddo - !$acc cache (a) ! { dg-error "inside of loop" } + !$acc cache (a(1:10)) ! { dg-error "ACC CACHE directive must be inside of loop" } do i = 1,10 - !$acc cache(a) + !$acc cache(a(i:i+1)) enddo do i = 1,10 + !$acc cache(a(i:i+1)) a(i) = i - !$acc cache(a) + !$acc cache(a(i+2:i+2+1)) enddo end subroutine test1 diff --git libgomp/ChangeLog libgomp/ChangeLog index f6d7dd2..5c7f41a 100644 --- libgomp/ChangeLog +++ libgomp/ChangeLog @@ -1,3 +1,10 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + * testsuite/libgomp.oacc-c-c++-common/cache-1.c: #include + "../../../gcc/testsuite/c-c++-common/goacc/cache-1.c". + * testsuite/libgomp.oacc-fortran/cache-1.f95: New file. + 2016-06-03 Chung-Lin Tang <cltang@codesourcery.com> * testsuite/libgomp.oacc-fortran/reduction-8.f90: New testcase. diff --git libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c index 3f1f0bb..16aaed5 100644 --- libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c +++ libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c @@ -1,48 +1,3 @@ -int -main (int argc, char **argv) -{ -#define N 2 - int a[N], b[N]; - int i; +/* OpenACC cache directive. */ - for (i = 0; i < N; i++) - { - a[i] = 3; - b[i] = 0; - } - -#pragma acc parallel copyin (a[0:N]) copyout (b[0:N]) -{ - int ii; - - for (ii = 0; ii < N; ii++) - { - const int idx = ii; - int n = 1; - const int len = n; - -#pragma acc cache (a[0:N]) - -#pragma acc cache (a[0:N], b[0:N]) - -#pragma acc cache (a[0]) - -#pragma acc cache (a[0], a[1], b[0:N]) - -#pragma acc cache (a[idx]) - -#pragma acc cache (a[idx:len]) - - b[ii] = a[ii]; - } -} - - - for (i = 0; i < N; i++) - { - if (a[i] != b[i]) - __builtin_abort (); - } - - return 0; -} +#include "../../../gcc/testsuite/c-c++-common/goacc/cache-1.c" diff --git libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 new file mode 100644 index 0000000..37313d8 --- /dev/null +++ libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 @@ -0,0 +1,6 @@ +! OpenACC cache directive. +! { dg-do run } +! { dg-additional-options "-std=f2008" } +! { dg-additional-options "-cpp" } + +#include "../../../gcc/testsuite/gfortran.dg/goacc/cache-1.f95" commit 11334cfa66bcb2a2096e02c3d5bf92e0068f3909 Author: tschwinge <tschwinge@138bc75d-0d04-0410-961f-82ee72b054a4> Date: Fri Jun 10 09:46:04 2016 +0000 [PR c/71381] C/C++ OpenACC cache directive rejects valid syntax Backport trunk r237290: gcc/c/ PR c/71381 * c-parser.c (c_parser_omp_variable_list) <OMP_CLAUSE__CACHE_>: Loosen checking. gcc/cp/ PR c/71381 * parser.c (cp_parser_omp_var_list_no_open) <OMP_CLAUSE__CACHE_>: Loosen checking. gcc/fortran/ PR c/71381 * openmp.c (gfc_match_oacc_cache): Add comment. gcc/testsuite/ PR c/71381 * c-c++-common/goacc/cache-1.c: Update. Move invalid usage tests to... * c-c++-common/goacc/cache-2.c: ... this new file. * gfortran.dg/goacc/cache-1.f95: Move invalid usage tests to... * gfortran.dg/goacc/cache-2.f95: ... this new file. * gfortran.dg/goacc/coarray.f95: Update OpenACC cache directive usage. * gfortran.dg/goacc/cray.f95: Likewise. * gfortran.dg/goacc/loop-1.f95: Likewise. libgomp/ PR c/71381 * testsuite/libgomp.oacc-c-c++-common/cache-1.c: #include "../../../gcc/testsuite/c-c++-common/goacc/cache-1.c". * testsuite/libgomp.oacc-fortran/cache-1.f95: New file. gcc/ * omp-low.c (scan_sharing_clauses): Don't expect OMP_CLAUSE__CACHE_. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/branches/gcc-6-branch@237295 138bc75d-0d04-0410-961f-82ee72b054a4 --- gcc/ChangeLog | 4 ++ gcc/c/ChangeLog | 7 +++ gcc/c/c-parser.c | 21 +------ gcc/cp/ChangeLog | 7 +++ gcc/cp/parser.c | 21 +------ gcc/fortran/ChangeLog | 6 ++ gcc/fortran/openmp.c | 4 ++ gcc/omp-low.c | 6 -- gcc/testsuite/ChangeLog | 14 +++++ gcc/testsuite/c-c++-common/goacc/cache-1.c | 66 ++++++++-------------- .../c-c++-common/goacc/{cache-1.c => cache-2.c} | 39 ++----------- gcc/testsuite/gfortran.dg/goacc/cache-1.f95 | 7 +-- gcc/testsuite/gfortran.dg/goacc/cache-2.f95 | 12 ++++ gcc/testsuite/gfortran.dg/goacc/coarray.f95 | 2 +- gcc/testsuite/gfortran.dg/goacc/cray.f95 | 3 +- gcc/testsuite/gfortran.dg/goacc/loop-1.f95 | 7 ++- libgomp/ChangeLog | 8 +++ .../testsuite/libgomp.oacc-c-c++-common/cache-1.c | 49 +--------------- libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 | 6 ++ 19 files changed, 110 insertions(+), 179 deletions(-) diff --git gcc/ChangeLog gcc/ChangeLog index 4b7b8f6..cbdcc42 100644 --- gcc/ChangeLog +++ gcc/ChangeLog @@ -1,5 +1,9 @@ 2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + Backport from trunk r237290: + * omp-low.c (scan_sharing_clauses): Don't expect + OMP_CLAUSE__CACHE_. + Backport trunk r235964: 2016-05-06 Nathan Sidwell <nathan@codesourcery.com> diff --git gcc/c/ChangeLog gcc/c/ChangeLog index fc89216..7edb380 100644 --- gcc/c/ChangeLog +++ gcc/c/ChangeLog @@ -1,3 +1,10 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + Backport from trunk r237290: + * c-parser.c (c_parser_omp_variable_list) <OMP_CLAUSE__CACHE_>: + Loosen checking. + 2016-05-30 Jakub Jelinek <jakub@redhat.com> PR c++/71349 diff --git gcc/c/c-parser.c gcc/c/c-parser.c index 77c49a1..c9eb8dd 100644 --- gcc/c/c-parser.c +++ gcc/c/c-parser.c @@ -10595,6 +10595,8 @@ c_parser_omp_variable_list (c_parser *parser, switch (kind) { case OMP_CLAUSE__CACHE_: + /* The OpenACC cache directive explicitly only allows "array + elements or subarrays". */ if (c_parser_peek_token (parser)->type != CPP_OPEN_SQUARE) { c_parser_error (parser, "expected %<[%>"); @@ -10657,25 +10659,6 @@ c_parser_omp_variable_list (c_parser *parser, break; } - if (kind == OMP_CLAUSE__CACHE_) - { - if (TREE_CODE (low_bound) != INTEGER_CST - && !TREE_READONLY (low_bound)) - { - error_at (clause_loc, - "%qD is not a constant", low_bound); - t = error_mark_node; - } - - if (TREE_CODE (length) != INTEGER_CST - && !TREE_READONLY (length)) - { - error_at (clause_loc, - "%qD is not a constant", length); - t = error_mark_node; - } - } - t = tree_cons (low_bound, length, t); } break; diff --git gcc/cp/ChangeLog gcc/cp/ChangeLog index a55f04e..8fb22e8 100644 --- gcc/cp/ChangeLog +++ gcc/cp/ChangeLog @@ -1,3 +1,10 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + Backport from trunk r237290: + * parser.c (cp_parser_omp_var_list_no_open) <OMP_CLAUSE__CACHE_>: + Loosen checking. + 2016-06-08 Jakub Jelinek <jakub@redhat.com> PR c++/71442 diff --git gcc/cp/parser.c gcc/cp/parser.c index f626a5d..f59b03a 100644 --- gcc/cp/parser.c +++ gcc/cp/parser.c @@ -29964,6 +29964,8 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, switch (kind) { case OMP_CLAUSE__CACHE_: + /* The OpenACC cache directive explicitly only allows "array + elements or subarrays". */ if (cp_lexer_peek_token (parser->lexer)->type != CPP_OPEN_SQUARE) { error_at (token->location, "expected %<[%>"); @@ -30015,25 +30017,6 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, RT_CLOSE_SQUARE)) goto skip_comma; - if (kind == OMP_CLAUSE__CACHE_) - { - if (TREE_CODE (low_bound) != INTEGER_CST - && !TREE_READONLY (low_bound)) - { - error_at (token->location, - "%qD is not a constant", low_bound); - decl = error_mark_node; - } - - if (TREE_CODE (length) != INTEGER_CST - && !TREE_READONLY (length)) - { - error_at (token->location, - "%qD is not a constant", length); - decl = error_mark_node; - } - } - decl = tree_cons (low_bound, length, decl); } break; diff --git gcc/fortran/ChangeLog gcc/fortran/ChangeLog index 5a21c9b..de67856 100644 --- gcc/fortran/ChangeLog +++ gcc/fortran/ChangeLog @@ -1,3 +1,9 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + Backport from trunk r237290: + * openmp.c (gfc_match_oacc_cache): Add comment. + 2016-06-05 Andre Vehreschild <vehre@gcc.gnu.org> PR fortran/69659 diff --git gcc/fortran/openmp.c gcc/fortran/openmp.c index 0dd1a92..be70318 100644 --- gcc/fortran/openmp.c +++ gcc/fortran/openmp.c @@ -1623,6 +1623,10 @@ match gfc_match_oacc_cache (void) { gfc_omp_clauses *c = gfc_get_omp_clauses (); + /* The OpenACC cache directive explicitly only allows "array elements or + subarrays", which we're currently not checking here. Either check this + after the call of gfc_match_omp_variable_list, or add something like a + only_sections variant next to its allow_sections parameter. */ match m = gfc_match_omp_variable_list (" (", &c->lists[OMP_LIST_CACHE], true, NULL, NULL, true); diff --git gcc/omp-low.c gcc/omp-low.c index 4ad2625..e570c22 100644 --- gcc/omp-low.c +++ gcc/omp-low.c @@ -2201,9 +2201,6 @@ scan_sharing_clauses (tree clauses, omp_context *ctx, case OMP_CLAUSE_DEVICE_RESIDENT: case OMP_CLAUSE__CACHE_: - sorry ("Clause not supported yet"); - break; - default: gcc_unreachable (); } @@ -2369,9 +2366,6 @@ scan_sharing_clauses (tree clauses, omp_context *ctx, case OMP_CLAUSE_DEVICE_RESIDENT: case OMP_CLAUSE__CACHE_: - sorry ("Clause not supported yet"); - break; - default: gcc_unreachable (); } diff --git gcc/testsuite/ChangeLog gcc/testsuite/ChangeLog index 8437f84..5e3f7e3 100644 --- gcc/testsuite/ChangeLog +++ gcc/testsuite/ChangeLog @@ -1,3 +1,17 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + Backport from trunk r237290: + * c-c++-common/goacc/cache-1.c: Update. Move invalid usage tests + to... + * c-c++-common/goacc/cache-2.c: ... this new file. + * gfortran.dg/goacc/cache-1.f95: Move invalid usage tests to... + * gfortran.dg/goacc/cache-2.f95: ... this new file. + * gfortran.dg/goacc/coarray.f95: Update OpenACC cache directive + usage. + * gfortran.dg/goacc/cray.f95: Likewise. + * gfortran.dg/goacc/loop-1.f95: Likewise. + 2016-06-09 Michael Meissner <meissner@linux.vnet.ibm.com> Back port from trunk diff --git gcc/testsuite/c-c++-common/goacc/cache-1.c gcc/testsuite/c-c++-common/goacc/cache-1.c index 9503341..1d4759e 100644 --- gcc/testsuite/c-c++-common/goacc/cache-1.c +++ gcc/testsuite/c-c++-common/goacc/cache-1.c @@ -1,3 +1,7 @@ +/* OpenACC cache directive: valid usage. */ +/* For execution testing, this file is "#include"d from + libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c. */ + int main (int argc, char **argv) { @@ -21,57 +25,31 @@ main (int argc, char **argv) int n = 1; const int len = n; -#pragma acc cache /* { dg-error "expected '\\$' before end of line" } */ - -#pragma acc cache a[0:N] /* { dg-error "expected '\\\(' before 'a'" } */ - /* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } 26 } */ - -#pragma acc cache (a) /* { dg-error "expected '\\\['" } */ - -#pragma acc cache ( /* { dg-error "expected (identifier|unqualified-id) before end of line" } */ - -#pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\$' token" } */ - -#pragma acc cache (,) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" } */ - -#pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - -#pragma acc cache (a[0:N],) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" "" { xfail c } } */ - -#pragma acc cache (a[0:N]) copyin (a[0:N]) /* { dg-error "expected end of line before 'copyin'" } */ - -#pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\\)' token" } */ - -#pragma acc cache (a[0:N] b[0:N]) /* { dg-error "expected '\\\)' before 'b'" } */ - -#pragma acc cache (a[0:N] b[0:N}) /* { dg-error "expected '\\\)' before 'b'" } */ - /* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } 47 } */ - -#pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - -#pragma acc cache (a[ii]) /* { dg-error "'ii' is not a constant" } */ - -#pragma acc cache (a[idx:n]) /* { dg-error "'n' is not a constant" } */ - -#pragma acc cache (a[0:N]) ( /* { dg-error "expected end of line before '\$' token" } */ - -#pragma acc cache (a[0:N]) ii /* { dg-error "expected end of line before 'ii'" } */ - -#pragma acc cache (a[0:N] ii) /* { dg-error "expected '\$' before 'ii'" } */ - + /* Have at it, GCC! */ #pragma acc cache (a[0:N]) - #pragma acc cache (a[0:N], a[0:N]) - #pragma acc cache (a[0:N], b[0:N]) - #pragma acc cache (a[0]) - #pragma acc cache (a[0], a[1], b[0:N]) - +#pragma acc cache (a[i - 5]) +#pragma acc cache (a[i + 5:len]) +#pragma acc cache (a[i + 5:len - 1]) +#pragma acc cache (b[i]) +#pragma acc cache (b[i:len]) +#pragma acc cache (a[ii]) +#pragma acc cache (a[ii:len]) +#pragma acc cache (b[ii - 1]) +#pragma acc cache (b[ii - 1:len]) +#pragma acc cache (b[i - ii + 1]) +#pragma acc cache (b[i + ii - 1:len]) +#pragma acc cache (b[i * ii - 1:len + 1]) +#pragma acc cache (a[idx + 2]) +#pragma acc cache (a[idx:len + 2]) #pragma acc cache (a[idx]) - #pragma acc cache (a[idx:len]) +#pragma acc cache (a[idx + 2:len]) +#pragma acc cache (a[idx + 2 + i:len]) +#pragma acc cache (a[idx + 2 + i + ii:len]) b[ii] = a[ii]; } diff --git gcc/testsuite/c-c++-common/goacc/cache-1.c gcc/testsuite/c-c++-common/goacc/cache-2.c similarity index 83% copy from gcc/testsuite/c-c++-common/goacc/cache-1.c copy to gcc/testsuite/c-c++-common/goacc/cache-2.c index 9503341..f717515 100644 --- gcc/testsuite/c-c++-common/goacc/cache-1.c +++ gcc/testsuite/c-c++-common/goacc/cache-2.c @@ -1,3 +1,5 @@ +/* OpenACC cache directive: invalid usage. */ + int main (int argc, char **argv) { @@ -22,57 +24,24 @@ main (int argc, char **argv) const int len = n; #pragma acc cache /* { dg-error "expected '\\$' before end of line" } */ - #pragma acc cache a[0:N] /* { dg-error "expected '\\\(' before 'a'" } */ - /* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } 26 } */ - + /* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } 27 } */ #pragma acc cache (a) /* { dg-error "expected '\\\['" } */ - #pragma acc cache ( /* { dg-error "expected (identifier|unqualified-id) before end of line" } */ - #pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\$' token" } */ - #pragma acc cache (,) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" } */ - #pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - #pragma acc cache (a[0:N],) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" "" { xfail c } } */ - #pragma acc cache (a[0:N]) copyin (a[0:N]) /* { dg-error "expected end of line before 'copyin'" } */ - #pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\\)' token" } */ - #pragma acc cache (a[0:N] b[0:N]) /* { dg-error "expected '\\\)' before 'b'" } */ - #pragma acc cache (a[0:N] b[0:N}) /* { dg-error "expected '\\\)' before 'b'" } */ - /* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } 47 } */ - + /* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } 38 } */ #pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - -#pragma acc cache (a[ii]) /* { dg-error "'ii' is not a constant" } */ - -#pragma acc cache (a[idx:n]) /* { dg-error "'n' is not a constant" } */ - #pragma acc cache (a[0:N]) ( /* { dg-error "expected end of line before '\$' token" } */ - #pragma acc cache (a[0:N]) ii /* { dg-error "expected end of line before 'ii'" } */ - #pragma acc cache (a[0:N] ii) /* { dg-error "expected '\$' before 'ii'" } */ -#pragma acc cache (a[0:N]) - -#pragma acc cache (a[0:N], a[0:N]) - -#pragma acc cache (a[0:N], b[0:N]) - -#pragma acc cache (a[0]) - -#pragma acc cache (a[0], a[1], b[0:N]) - -#pragma acc cache (a[idx]) - -#pragma acc cache (a[idx:len]) - b[ii] = a[ii]; } } diff --git gcc/testsuite/gfortran.dg/goacc/cache-1.f95 gcc/testsuite/gfortran.dg/goacc/cache-1.f95 index 2aa9e05..39fbf2c 100644 --- gcc/testsuite/gfortran.dg/goacc/cache-1.f95 +++ gcc/testsuite/gfortran.dg/goacc/cache-1.f95 @@ -1,4 +1,6 @@ -! { dg-do compile } +! OpenACC cache directive: valid usage. +! For execution testing, this file is "#include"d from +! libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95. ! { dg-additional-options "-std=f2008" } program test @@ -6,11 +8,8 @@ program test integer :: i, d(10), e(5,13) do concurrent (i=1:5) - !$acc cache (d) !$acc cache (d(1:3)) !$acc cache (d(i:i+2)) - - !$acc cache (e) !$acc cache (e(1:3,2:4)) !$acc cache (e(i:i+2,i+1:i+3)) enddo diff --git gcc/testsuite/gfortran.dg/goacc/cache-2.f95 gcc/testsuite/gfortran.dg/goacc/cache-2.f95 new file mode 100644 index 0000000..be81878 --- /dev/null +++ gcc/testsuite/gfortran.dg/goacc/cache-2.f95 @@ -0,0 +1,12 @@ +! OpenACC cache directive: invalid usage. +! { dg-additional-options "-std=f2008" } + +program test + implicit none + integer :: i, d(10), e(5,13) + + do concurrent (i=1:5) + !$acc cache (d) ! { dg-error "" "TODO" { xfail *-*-* } } + !$acc cache (e) ! { dg-error "" "TODO" { xfail *-*-* } } + enddo +end diff --git gcc/testsuite/gfortran.dg/goacc/coarray.f95 gcc/testsuite/gfortran.dg/goacc/coarray.f95 index 932e1f7..f30917b8 100644 --- gcc/testsuite/gfortran.dg/goacc/coarray.f95 +++ gcc/testsuite/gfortran.dg/goacc/coarray.f95 @@ -24,7 +24,7 @@ contains !$acc end parallel loop !$acc parallel loop do i = 1,5 - !$acc cache (a) + !$acc cache (a) ! { dg-error "" "TODO" { xfail *-*-* } } enddo !$acc end parallel loop !$acc update device (a) diff --git gcc/testsuite/gfortran.dg/goacc/cray.f95 gcc/testsuite/gfortran.dg/goacc/cray.f95 index a35ab0d..705c18c 100644 --- gcc/testsuite/gfortran.dg/goacc/cray.f95 +++ gcc/testsuite/gfortran.dg/goacc/cray.f95 @@ -44,7 +44,8 @@ contains !$acc end parallel loop !$acc parallel loop do i = 1,5 - !$acc cache (ptr) ! TODO: This must fail, as in openacc-1_0-branch + !TODO: This must fail, as in openacc-1_0-branch. + !$acc cache (ptr) ! { dg-error "" "TODO" { xfail *-*-* } } enddo !$acc end parallel loop !$acc update device (ptr) diff --git gcc/testsuite/gfortran.dg/goacc/loop-1.f95 gcc/testsuite/gfortran.dg/goacc/loop-1.f95 index b5f9e03..a605f03 100644 --- gcc/testsuite/gfortran.dg/goacc/loop-1.f95 +++ gcc/testsuite/gfortran.dg/goacc/loop-1.f95 @@ -158,15 +158,16 @@ subroutine test1 enddo - !$acc cache (a) ! { dg-error "inside of loop" } + !$acc cache (a(1:10)) ! { dg-error "ACC CACHE directive must be inside of loop" } do i = 1,10 - !$acc cache(a) + !$acc cache(a(i:i+1)) enddo do i = 1,10 + !$acc cache(a(i:i+1)) a(i) = i - !$acc cache(a) + !$acc cache(a(i+2:i+2+1)) enddo end subroutine test1 diff --git libgomp/ChangeLog libgomp/ChangeLog index 7b124eb..f6da5c4 100644 --- libgomp/ChangeLog +++ libgomp/ChangeLog @@ -1,3 +1,11 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + Backport from trunk r237290: + * testsuite/libgomp.oacc-c-c++-common/cache-1.c: #include + "../../../gcc/testsuite/c-c++-common/goacc/cache-1.c". + * testsuite/libgomp.oacc-fortran/cache-1.f95: New file. + 2016-05-23 Martin Jambor <mjambor@suse.cz> * testsuite/libgomp.hsa.c/switch-sbr-2.c: New test. diff --git libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c index 3f1f0bb..16aaed5 100644 --- libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c +++ libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c @@ -1,48 +1,3 @@ -int -main (int argc, char **argv) -{ -#define N 2 - int a[N], b[N]; - int i; +/* OpenACC cache directive. */ - for (i = 0; i < N; i++) - { - a[i] = 3; - b[i] = 0; - } - -#pragma acc parallel copyin (a[0:N]) copyout (b[0:N]) -{ - int ii; - - for (ii = 0; ii < N; ii++) - { - const int idx = ii; - int n = 1; - const int len = n; - -#pragma acc cache (a[0:N]) - -#pragma acc cache (a[0:N], b[0:N]) - -#pragma acc cache (a[0]) - -#pragma acc cache (a[0], a[1], b[0:N]) - -#pragma acc cache (a[idx]) - -#pragma acc cache (a[idx:len]) - - b[ii] = a[ii]; - } -} - - - for (i = 0; i < N; i++) - { - if (a[i] != b[i]) - __builtin_abort (); - } - - return 0; -} +#include "../../../gcc/testsuite/c-c++-common/goacc/cache-1.c" diff --git libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 new file mode 100644 index 0000000..37313d8 --- /dev/null +++ libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 @@ -0,0 +1,6 @@ +! OpenACC cache directive. +! { dg-do run } +! { dg-additional-options "-std=f2008" } +! { dg-additional-options "-cpp" } + +#include "../../../gcc/testsuite/gfortran.dg/goacc/cache-1.f95" commit ca3738135ba830829657e20b66f197616bbf733a Author: tschwinge <tschwinge@138bc75d-0d04-0410-961f-82ee72b054a4> Date: Fri Jun 10 10:12:22 2016 +0000 [PR c/71381] C/C++ OpenACC cache directive rejects valid syntax libgomp/ PR c/71381 * testsuite/libgomp.oacc-fortran/cache-1.f90: Remove file. Backport trunk r237290: gcc/c/ PR c/71381 * c-parser.c (c_parser_omp_variable_list) <OMP_CLAUSE__CACHE_>: Loosen checking. gcc/cp/ PR c/71381 * parser.c (cp_parser_omp_var_list_no_open) <OMP_CLAUSE__CACHE_>: Loosen checking. gcc/fortran/ PR c/71381 * openmp.c (gfc_match_oacc_cache): Add comment. gcc/testsuite/ PR c/71381 * c-c++-common/goacc/cache-1.c: Update. Move invalid usage tests to... * c-c++-common/goacc/cache-2.c: ... this new file. * gfortran.dg/goacc/cache-1.f95: Move invalid usage tests to... * gfortran.dg/goacc/cache-2.f95: ... this new file. * gfortran.dg/goacc/coarray.f95: Update OpenACC cache directive usage. * gfortran.dg/goacc/cray.f95: Likewise. * gfortran.dg/goacc/loop-1.f95: Likewise. libgomp/ PR c/71381 * testsuite/libgomp.oacc-c-c++-common/cache-1.c: #include "../../../gcc/testsuite/c-c++-common/goacc/cache-1.c". * testsuite/libgomp.oacc-fortran/cache-1.f95: New file. gcc/ * omp-low.c (scan_sharing_clauses): Don't expect OMP_CLAUSE__CACHE_. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/branches/gomp-4_0-branch@237299 138bc75d-0d04-0410-961f-82ee72b054a4 --- gcc/ChangeLog.gomp | 6 ++ gcc/c/ChangeLog.gomp | 7 +++ gcc/c/c-parser.c | 21 +------ gcc/cp/ChangeLog.gomp | 7 +++ gcc/cp/parser.c | 21 +------ gcc/fortran/ChangeLog.gomp | 6 ++ gcc/fortran/openmp.c | 4 ++ gcc/omp-low.c | 14 ++--- gcc/testsuite/ChangeLog.gomp | 14 +++++ gcc/testsuite/c-c++-common/goacc/cache-1.c | 66 ++++++++-------------- .../c-c++-common/goacc/{cache-1.c => cache-2.c} | 39 ++----------- gcc/testsuite/gfortran.dg/goacc/cache-1.f95 | 7 +-- gcc/testsuite/gfortran.dg/goacc/cache-2.f95 | 12 ++++ gcc/testsuite/gfortran.dg/goacc/coarray.f95 | 2 +- gcc/testsuite/gfortran.dg/goacc/cray.f95 | 3 +- gcc/testsuite/gfortran.dg/goacc/loop-1.f95 | 7 ++- libgomp/ChangeLog.gomp | 9 +++ .../testsuite/libgomp.oacc-c-c++-common/cache-1.c | 49 +--------------- libgomp/testsuite/libgomp.oacc-fortran/cache-1.f90 | 30 ---------- libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 | 6 ++ 20 files changed, 117 insertions(+), 213 deletions(-) diff --git gcc/ChangeLog.gomp gcc/ChangeLog.gomp index a73eff6..4477abf 100644 --- gcc/ChangeLog.gomp +++ gcc/ChangeLog.gomp @@ -1,3 +1,9 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + Backport from trunk r237290: + * omp-low.c (scan_sharing_clauses): Don't expect + OMP_CLAUSE__CACHE_. + 2016-05-27 Cesar Philippidis <cesar@codesourcery.com> Backport trunk r236678: diff --git gcc/c/ChangeLog.gomp gcc/c/ChangeLog.gomp index 26de504..aa2ba02 100644 --- gcc/c/ChangeLog.gomp +++ gcc/c/ChangeLog.gomp @@ -1,3 +1,10 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + Backport from trunk r237290: + * c-parser.c (c_parser_omp_variable_list) <OMP_CLAUSE__CACHE_>: + Loosen checking. + 2016-06-08 Chung-Lin Tang <cltang@codesourcery.com> Backport trunk r237070: diff --git gcc/c/c-parser.c gcc/c/c-parser.c index 2f1c826..20b05d3 100644 --- gcc/c/c-parser.c +++ gcc/c/c-parser.c @@ -10634,6 +10634,8 @@ c_parser_omp_variable_list (c_parser *parser, switch (kind) { case OMP_CLAUSE__CACHE_: + /* The OpenACC cache directive explicitly only allows "array + elements or subarrays". */ if (c_parser_peek_token (parser)->type != CPP_OPEN_SQUARE) { c_parser_error (parser, "expected %<[%>"); @@ -10696,25 +10698,6 @@ c_parser_omp_variable_list (c_parser *parser, break; } - if (kind == OMP_CLAUSE__CACHE_) - { - if (TREE_CODE (low_bound) != INTEGER_CST - && !TREE_READONLY (low_bound)) - { - error_at (clause_loc, - "%qD is not a constant", low_bound); - t = error_mark_node; - } - - if (TREE_CODE (length) != INTEGER_CST - && !TREE_READONLY (length)) - { - error_at (clause_loc, - "%qD is not a constant", length); - t = error_mark_node; - } - } - t = tree_cons (low_bound, length, t); } break; diff --git gcc/cp/ChangeLog.gomp gcc/cp/ChangeLog.gomp index b3fef78..f030a59 100644 --- gcc/cp/ChangeLog.gomp +++ gcc/cp/ChangeLog.gomp @@ -1,3 +1,10 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + Backport from trunk r237290: + * parser.c (cp_parser_omp_var_list_no_open) <OMP_CLAUSE__CACHE_>: + Loosen checking. + 2016-06-08 Chung-Lin Tang <cltang@codesourcery.com> Backport trunk r237070: diff --git gcc/cp/parser.c gcc/cp/parser.c index 599ca77..53c7892 100644 --- gcc/cp/parser.c +++ gcc/cp/parser.c @@ -29960,6 +29960,8 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, switch (kind) { case OMP_CLAUSE__CACHE_: + /* The OpenACC cache directive explicitly only allows "array + elements or subarrays". */ if (cp_lexer_peek_token (parser->lexer)->type != CPP_OPEN_SQUARE) { error_at (token->location, "expected %<[%>"); @@ -30011,25 +30013,6 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, RT_CLOSE_SQUARE)) goto skip_comma; - if (kind == OMP_CLAUSE__CACHE_) - { - if (TREE_CODE (low_bound) != INTEGER_CST - && !TREE_READONLY (low_bound)) - { - error_at (token->location, - "%qD is not a constant", low_bound); - decl = error_mark_node; - } - - if (TREE_CODE (length) != INTEGER_CST - && !TREE_READONLY (length)) - { - error_at (token->location, - "%qD is not a constant", length); - decl = error_mark_node; - } - } - decl = tree_cons (low_bound, length, decl); } break; diff --git gcc/fortran/ChangeLog.gomp gcc/fortran/ChangeLog.gomp index 24fec48..527d088 100644 --- gcc/fortran/ChangeLog.gomp +++ gcc/fortran/ChangeLog.gomp @@ -1,3 +1,9 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + Backport from trunk r237290: + * openmp.c (gfc_match_oacc_cache): Add comment. + 2016-06-08 Chung-Lin Tang <cltang@codesourcery.com> Backport trunk r237070: diff --git gcc/fortran/openmp.c gcc/fortran/openmp.c index a2a0e4b..cc4583d 100644 --- gcc/fortran/openmp.c +++ gcc/fortran/openmp.c @@ -1799,6 +1799,10 @@ match gfc_match_oacc_cache (void) { gfc_omp_clauses *c = gfc_get_omp_clauses (); + /* The OpenACC cache directive explicitly only allows "array elements or + subarrays", which we're currently not checking here. Either check this + after the call of gfc_match_omp_variable_list, or add something like a + only_sections variant next to its allow_sections parameter. */ match m = gfc_match_omp_variable_list (" (", &c->lists[OMP_LIST_CACHE], true, NULL, NULL, true); diff --git gcc/omp-low.c gcc/omp-low.c index 200d331..40ac8c8 100644 --- gcc/omp-low.c +++ gcc/omp-low.c @@ -2232,13 +2232,10 @@ scan_sharing_clauses (tree clauses, omp_context *ctx, install_var_local (decl, ctx); break; - case OMP_CLAUSE_DEVICE_RESIDENT: - case OMP_CLAUSE__CACHE_: - sorry ("Clause not supported yet"); - break; - case OMP_CLAUSE_BIND: + case OMP_CLAUSE_DEVICE_RESIDENT: case OMP_CLAUSE_NOHOST: + case OMP_CLAUSE__CACHE_: default: gcc_unreachable (); } @@ -2403,13 +2400,10 @@ scan_sharing_clauses (tree clauses, omp_context *ctx, case OMP_CLAUSE_DEVICE_TYPE: break; - case OMP_CLAUSE_DEVICE_RESIDENT: - case OMP_CLAUSE__CACHE_: - sorry ("Clause not supported yet"); - break; - case OMP_CLAUSE_BIND: + case OMP_CLAUSE_DEVICE_RESIDENT: case OMP_CLAUSE_NOHOST: + case OMP_CLAUSE__CACHE_: default: gcc_unreachable (); } diff --git gcc/testsuite/ChangeLog.gomp gcc/testsuite/ChangeLog.gomp index 41669bc..eef9425 100644 --- gcc/testsuite/ChangeLog.gomp +++ gcc/testsuite/ChangeLog.gomp @@ -1,3 +1,17 @@ +2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + + PR c/71381 + Backport from trunk r237290: + * c-c++-common/goacc/cache-1.c: Update. Move invalid usage tests + to... + * c-c++-common/goacc/cache-2.c: ... this new file. + * gfortran.dg/goacc/cache-1.f95: Move invalid usage tests to... + * gfortran.dg/goacc/cache-2.f95: ... this new file. + * gfortran.dg/goacc/coarray.f95: Update OpenACC cache directive + usage. + * gfortran.dg/goacc/cray.f95: Likewise. + * gfortran.dg/goacc/loop-1.f95: Likewise. + 2016-05-27 Cesar Philippidis <cesar@codesourcery.com> * c-c++-common/goacc/kernels-loop-offload-alias-none.c: Add xfails. diff --git gcc/testsuite/c-c++-common/goacc/cache-1.c gcc/testsuite/c-c++-common/goacc/cache-1.c index 9503341..1d4759e 100644 --- gcc/testsuite/c-c++-common/goacc/cache-1.c +++ gcc/testsuite/c-c++-common/goacc/cache-1.c @@ -1,3 +1,7 @@ +/* OpenACC cache directive: valid usage. */ +/* For execution testing, this file is "#include"d from + libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c. */ + int main (int argc, char **argv) { @@ -21,57 +25,31 @@ main (int argc, char **argv) int n = 1; const int len = n; -#pragma acc cache /* { dg-error "expected '\\$' before end of line" } */ - -#pragma acc cache a[0:N] /* { dg-error "expected '\\\(' before 'a'" } */ - /* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } 26 } */ - -#pragma acc cache (a) /* { dg-error "expected '\\\['" } */ - -#pragma acc cache ( /* { dg-error "expected (identifier|unqualified-id) before end of line" } */ - -#pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\$' token" } */ - -#pragma acc cache (,) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" } */ - -#pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - -#pragma acc cache (a[0:N],) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" "" { xfail c } } */ - -#pragma acc cache (a[0:N]) copyin (a[0:N]) /* { dg-error "expected end of line before 'copyin'" } */ - -#pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\\)' token" } */ - -#pragma acc cache (a[0:N] b[0:N]) /* { dg-error "expected '\\\)' before 'b'" } */ - -#pragma acc cache (a[0:N] b[0:N}) /* { dg-error "expected '\\\)' before 'b'" } */ - /* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } 47 } */ - -#pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - -#pragma acc cache (a[ii]) /* { dg-error "'ii' is not a constant" } */ - -#pragma acc cache (a[idx:n]) /* { dg-error "'n' is not a constant" } */ - -#pragma acc cache (a[0:N]) ( /* { dg-error "expected end of line before '\$' token" } */ - -#pragma acc cache (a[0:N]) ii /* { dg-error "expected end of line before 'ii'" } */ - -#pragma acc cache (a[0:N] ii) /* { dg-error "expected '\$' before 'ii'" } */ - + /* Have at it, GCC! */ #pragma acc cache (a[0:N]) - #pragma acc cache (a[0:N], a[0:N]) - #pragma acc cache (a[0:N], b[0:N]) - #pragma acc cache (a[0]) - #pragma acc cache (a[0], a[1], b[0:N]) - +#pragma acc cache (a[i - 5]) +#pragma acc cache (a[i + 5:len]) +#pragma acc cache (a[i + 5:len - 1]) +#pragma acc cache (b[i]) +#pragma acc cache (b[i:len]) +#pragma acc cache (a[ii]) +#pragma acc cache (a[ii:len]) +#pragma acc cache (b[ii - 1]) +#pragma acc cache (b[ii - 1:len]) +#pragma acc cache (b[i - ii + 1]) +#pragma acc cache (b[i + ii - 1:len]) +#pragma acc cache (b[i * ii - 1:len + 1]) +#pragma acc cache (a[idx + 2]) +#pragma acc cache (a[idx:len + 2]) #pragma acc cache (a[idx]) - #pragma acc cache (a[idx:len]) +#pragma acc cache (a[idx + 2:len]) +#pragma acc cache (a[idx + 2 + i:len]) +#pragma acc cache (a[idx + 2 + i + ii:len]) b[ii] = a[ii]; } diff --git gcc/testsuite/c-c++-common/goacc/cache-1.c gcc/testsuite/c-c++-common/goacc/cache-2.c similarity index 83% copy from gcc/testsuite/c-c++-common/goacc/cache-1.c copy to gcc/testsuite/c-c++-common/goacc/cache-2.c index 9503341..f717515 100644 --- gcc/testsuite/c-c++-common/goacc/cache-1.c +++ gcc/testsuite/c-c++-common/goacc/cache-2.c @@ -1,3 +1,5 @@ +/* OpenACC cache directive: invalid usage. */ + int main (int argc, char **argv) { @@ -22,57 +24,24 @@ main (int argc, char **argv) const int len = n; #pragma acc cache /* { dg-error "expected '\\$' before end of line" } */ - #pragma acc cache a[0:N] /* { dg-error "expected '\\\(' before 'a'" } */ - /* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } 26 } */ - + /* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } 27 } */ #pragma acc cache (a) /* { dg-error "expected '\\\['" } */ - #pragma acc cache ( /* { dg-error "expected (identifier|unqualified-id) before end of line" } */ - #pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\$' token" } */ - #pragma acc cache (,) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" } */ - #pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - #pragma acc cache (a[0:N],) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" "" { xfail c } } */ - #pragma acc cache (a[0:N]) copyin (a[0:N]) /* { dg-error "expected end of line before 'copyin'" } */ - #pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\\)' token" } */ - #pragma acc cache (a[0:N] b[0:N]) /* { dg-error "expected '\\\)' before 'b'" } */ - #pragma acc cache (a[0:N] b[0:N}) /* { dg-error "expected '\\\)' before 'b'" } */ - /* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } 47 } */ - + /* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } 38 } */ #pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */ - -#pragma acc cache (a[ii]) /* { dg-error "'ii' is not a constant" } */ - -#pragma acc cache (a[idx:n]) /* { dg-error "'n' is not a constant" } */ - #pragma acc cache (a[0:N]) ( /* { dg-error "expected end of line before '\$' token" } */ - #pragma acc cache (a[0:N]) ii /* { dg-error "expected end of line before 'ii'" } */ - #pragma acc cache (a[0:N] ii) /* { dg-error "expected '\$' before 'ii'" } */ -#pragma acc cache (a[0:N]) - -#pragma acc cache (a[0:N], a[0:N]) - -#pragma acc cache (a[0:N], b[0:N]) - -#pragma acc cache (a[0]) - -#pragma acc cache (a[0], a[1], b[0:N]) - -#pragma acc cache (a[idx]) - -#pragma acc cache (a[idx:len]) - b[ii] = a[ii]; } } diff --git gcc/testsuite/gfortran.dg/goacc/cache-1.f95 gcc/testsuite/gfortran.dg/goacc/cache-1.f95 index 2aa9e05..39fbf2c 100644 --- gcc/testsuite/gfortran.dg/goacc/cache-1.f95 +++ gcc/testsuite/gfortran.dg/goacc/cache-1.f95 @@ -1,4 +1,6 @@ -! { dg-do compile } +! OpenACC cache directive: valid usage. +! For execution testing, this file is "#include"d from +! libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95. ! { dg-additional-options "-std=f2008" } program test @@ -6,11 +8,8 @@ program test integer :: i, d(10), e(5,13) do concurrent (i=1:5) - !$acc cache (d) !$acc cache (d(1:3)) !$acc cache (d(i:i+2)) - - !$acc cache (e) !$acc cache (e(1:3,2:4)) !$acc cache (e(i:i+2,i+1:i+3)) enddo diff --git gcc/testsuite/gfortran.dg/goacc/cache-2.f95 gcc/testsuite/gfortran.dg/goacc/cache-2.f95 new file mode 100644 index 0000000..be81878 --- /dev/null +++ gcc/testsuite/gfortran.dg/goacc/cache-2.f95 @@ -0,0 +1,12 @@ +! OpenACC cache directive: invalid usage. +! { dg-additional-options "-std=f2008" } + +program test + implicit none + integer :: i, d(10), e(5,13) + + do concurrent (i=1:5) + !$acc cache (d) ! { dg-error "" "TODO" { xfail *-*-* } } + !$acc cache (e) ! { dg-error "" "TODO" { xfail *-*-* } } + enddo +end diff --git gcc/testsuite/gfortran.dg/goacc/coarray.f95 gcc/testsuite/gfortran.dg/goacc/coarray.f95 index 932e1f7..f30917b8 100644 --- gcc/testsuite/gfortran.dg/goacc/coarray.f95 +++ gcc/testsuite/gfortran.dg/goacc/coarray.f95 @@ -24,7 +24,7 @@ contains !$acc end parallel loop !$acc parallel loop do i = 1,5 - !$acc cache (a) + !$acc cache (a) ! { dg-error "" "TODO" { xfail *-*-* } } enddo !$acc end parallel loop !$acc update device (a) diff --git gcc/testsuite/gfortran.dg/goacc/cray.f95 gcc/testsuite/gfortran.dg/goacc/cray.f95 index a35ab0d..705c18c 100644 --- gcc/testsuite/gfortran.dg/goacc/cray.f95 +++ gcc/testsuite/gfortran.dg/goacc/cray.f95 @@ -44,7 +44,8 @@ contains !$acc end parallel loop !$acc parallel loop do i = 1,5 - !$acc cache (ptr) ! TODO: This must fail, as in openacc-1_0-branch + !TODO: This must fail, as in openacc-1_0-branch. + !$acc cache (ptr) ! { dg-error "" "TODO" { xfail *-*-* } } enddo !$acc end parallel loop !$acc update device (ptr) diff --git gcc/testsuite/gfortran.dg/goacc/loop-1.f95 gcc/testsuite/gfortran.dg/goacc/loop-1.f95 index b5f9e03..a605f03 100644 --- gcc/testsuite/gfortran.dg/goacc/loop-1.f95 +++ gcc/testsuite/gfortran.dg/goacc/loop-1.f95 @@ -158,15 +158,16 @@ subroutine test1 enddo - !$acc cache (a) ! { dg-error "inside of loop" } + !$acc cache (a(1:10)) ! { dg-error "ACC CACHE directive must be inside of loop" } do i = 1,10 - !$acc cache(a) + !$acc cache(a(i:i+1)) enddo do i = 1,10 + !$acc cache(a(i:i+1)) a(i) = i - !$acc cache(a) + !$acc cache(a(i+2:i+2+1)) enddo end subroutine test1 diff --git libgomp/ChangeLog.gomp libgomp/ChangeLog.gomp index ccee4d1c..17c26c5 100644 --- libgomp/ChangeLog.gomp +++ libgomp/ChangeLog.gomp @@ -1,5 +1,14 @@ 2016-06-10 Thomas Schwinge <thomas@codesourcery.com> + PR c/71381 + * testsuite/libgomp.oacc-fortran/cache-1.f90: Remove file. + + PR c/71381 + Backport from trunk r237290: + * testsuite/libgomp.oacc-c-c++-common/cache-1.c: #include + "../../../gcc/testsuite/c-c++-common/goacc/cache-1.c". + * testsuite/libgomp.oacc-fortran/cache-1.f95: New file. + * testsuite/libgomp.oacc-c++/template-reduction.C: Remove XFAIL. 2016-06-08 Chung-Lin Tang <cltang@codesourcery.com> diff --git libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c index 3f1f0bb..16aaed5 100644 --- libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c +++ libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c @@ -1,48 +1,3 @@ -int -main (int argc, char **argv) -{ -#define N 2 - int a[N], b[N]; - int i; +/* OpenACC cache directive. */ - for (i = 0; i < N; i++) - { - a[i] = 3; - b[i] = 0; - } - -#pragma acc parallel copyin (a[0:N]) copyout (b[0:N]) -{ - int ii; - - for (ii = 0; ii < N; ii++) - { - const int idx = ii; - int n = 1; - const int len = n; - -#pragma acc cache (a[0:N]) - -#pragma acc cache (a[0:N], b[0:N]) - -#pragma acc cache (a[0]) - -#pragma acc cache (a[0], a[1], b[0:N]) - -#pragma acc cache (a[idx]) - -#pragma acc cache (a[idx:len]) - - b[ii] = a[ii]; - } -} - - - for (i = 0; i < N; i++) - { - if (a[i] != b[i]) - __builtin_abort (); - } - - return 0; -} +#include "../../../gcc/testsuite/c-c++-common/goacc/cache-1.c" diff --git libgomp/testsuite/libgomp.oacc-fortran/cache-1.f90 libgomp/testsuite/libgomp.oacc-fortran/cache-1.f90 deleted file mode 100644 index 2aeacb9..0000000 --- libgomp/testsuite/libgomp.oacc-fortran/cache-1.f90 +++ /dev/null @@ -1,30 +0,0 @@ -! { dg-do run } - -program main - integer, parameter :: N = 8 - integer, dimension (N) :: a, b - integer :: i - integer :: idx, len - - idx = 1 - len = 2 - - !$acc parallel copyin (a(1:N)) copyout (b(1:N)) - do i = 1, N - - !$acc cache (a) - !$acc cache (a, b) - !$acc cache (a(2)) - !$acc cache (a(3), b(4)) - !$acc cache (a(1:N)) - !$acc cache (a(1:N), b(1:N)) - !$acc cache (a(idx)) - !$acc cache (a(idx:len), b(idx:len)) - !$acc cache (a(i:i+2)) - !$acc cache (a(i:i+2), b(i+1:i+3)) - - b(i) = a(i) - end do - !$acc end parallel - -end program diff --git libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 new file mode 100644 index 0000000..37313d8 --- /dev/null +++ libgomp/testsuite/libgomp.oacc-fortran/cache-1.f95 @@ -0,0 +1,6 @@ +! OpenACC cache directive. +! { dg-do run } +! { dg-additional-options "-std=f2008" } +! { dg-additional-options "-cpp" } + +#include "../../../gcc/testsuite/gfortran.dg/goacc/cache-1.f95"

[PING,PR,c/71381] C/C++ OpenACC cache directive rejects valid syntax

Commit Message

Comments

Patch