From mboxrd@z Thu Jan 1 00:00:00 1970 Return-Path: Received: (qmail 88827 invoked by alias); 24 Jun 2015 15:23:58 -0000 Mailing-List: contact gcc-patches-help@gcc.gnu.org; run by ezmlm Precedence: bulk List-Id: List-Archive: List-Post: List-Help: Sender: gcc-patches-owner@gcc.gnu.org Received: (qmail 88812 invoked by uid 89); 24 Jun 2015 15:23:58 -0000 Authentication-Results: sourceware.org; auth=none X-Virus-Found: No X-Spam-SWARE-Status: No, score=-1.5 required=5.0 tests=AWL,BAYES_00,RCVD_IN_DNSWL_LOW,SPF_PASS,T_FROM_12LTRDOM autolearn=ham version=3.3.2 X-HELO: relay1.mentorg.com Received: from relay1.mentorg.com (HELO relay1.mentorg.com) (192.94.38.131) by sourceware.org (qpsmtpd/0.93/v0.84-503-g423c35a) with ESMTP; Wed, 24 Jun 2015 15:23:56 +0000 Received: from svr-orw-fem-04.mgc.mentorg.com ([147.34.97.41]) by relay1.mentorg.com with esmtp id 1Z7mWn-0000CH-61 from James_Norris@mentor.com for gcc-patches@gcc.gnu.org; Wed, 24 Jun 2015 08:23:53 -0700 Received: from [172.30.80.204] (147.34.91.1) by svr-orw-fem-04.mgc.mentorg.com (147.34.97.41) with Microsoft SMTP Server id 14.3.224.2; Wed, 24 Jun 2015 08:23:52 -0700 Message-ID: <558ACB86.3020608@codesourcery.com> Date: Wed, 24 Jun 2015 15:33:00 -0000 From: James Norris User-Agent: Mozilla/5.0 (X11; Linux x86_64; rv:31.0) Gecko/20100101 Thunderbird/31.7.0 MIME-Version: 1.0 To: CC: Thomas Schwinge Subject: [gomp4] Additional tests for declare directive and fixes. Content-Type: multipart/mixed; boundary="------------060402060609040808080408" X-SW-Source: 2015-06/txt/msg01715.txt.bz2 --------------060402060609040808080408 Content-Type: text/plain; charset="utf-8"; format=flowed Content-Transfer-Encoding: 7bit Content-length: 161 Hi! The following patch adds additional testing of the declare directive and fixes for issues that arose from the testing. Committed to gomp-4_0-branch. Jim --------------060402060609040808080408 Content-Type: text/x-patch; name="declare.patch" Content-Transfer-Encoding: 7bit Content-Disposition: attachment; filename="declare.patch" Content-length: 5405 diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c index e7df751..bcbd163 100644 --- a/gcc/c/c-parser.c +++ b/gcc/c/c-parser.c @@ -1767,12 +1767,15 @@ finish_oacc_declare (tree fnbody, tree decls) break; } - stmt = make_node (OACC_DECLARE); - TREE_TYPE (stmt) = void_type_node; - OACC_DECLARE_CLAUSES (stmt) = ret_clauses; - SET_EXPR_LOCATION (stmt, loc); + if (ret_clauses) + { + stmt = make_node (OACC_DECLARE); + TREE_TYPE (stmt) = void_type_node; + OACC_DECLARE_CLAUSES (stmt) = ret_clauses; + SET_EXPR_LOCATION (stmt, loc); - tsi_link_before (&i, stmt, TSI_CONTINUE_LINKING); + tsi_link_before (&i, stmt, TSI_CONTINUE_LINKING); + } DECL_ATTRIBUTES (fndecl) = remove_attribute ("oacc declare", DECL_ATTRIBUTES (fndecl)); @@ -12812,6 +12815,14 @@ c_parser_oacc_declare (c_parser *parser) error = true; continue; } + else if (TREE_PUBLIC (decl)) + { + error_at (loc, + "invalid use of % variable %qD " + "in %<#pragma acc declare%>", decl); + error = true; + continue; + } break; } diff --git a/gcc/cp/decl.c b/gcc/cp/decl.c index 15da51e..a35f599 100644 --- a/gcc/cp/decl.c +++ b/gcc/cp/decl.c @@ -14343,7 +14343,17 @@ finish_oacc_declare (tree fndecl, tree decls) { t = tsi_stmt (i); if (TREE_CODE (t) == BIND_EXPR) - list = BIND_EXPR_BODY (t); + { + list = BIND_EXPR_BODY (t); + if (TREE_CODE (list) != STATEMENT_LIST) + { + stmt = list; + list = alloc_stmt_list (); + BIND_EXPR_BODY (t) = list; + i = tsi_start (list); + tsi_link_after (&i, stmt, TSI_CONTINUE_LINKING); + } + } } if (clauses) @@ -14371,11 +14381,11 @@ finish_oacc_declare (tree fndecl, tree decls) } } - if (!found) - { - i = tsi_start (list); - tsi_link_before (&i, stmt, TSI_CONTINUE_LINKING); - } + if (!found) + { + i = tsi_start (list); + tsi_link_before (&i, stmt, TSI_CONTINUE_LINKING); + } } while (oacc_returns) @@ -14405,18 +14415,21 @@ finish_oacc_declare (tree fndecl, tree decls) free (r); } - for (i = tsi_start (list); !tsi_end_p (i); tsi_next (&i)) + if (ret_clauses) { - if (tsi_end_p (i)) - break; - } + for (i = tsi_start (list); !tsi_end_p (i); tsi_next (&i)) + { + if (tsi_end_p (i)) + break; + } - stmt = make_node (OACC_DECLARE); - TREE_TYPE (stmt) = void_type_node; - OMP_STANDALONE_CLAUSES (stmt) = ret_clauses; - SET_EXPR_LOCATION (stmt, loc); + stmt = make_node (OACC_DECLARE); + TREE_TYPE (stmt) = void_type_node; + OMP_STANDALONE_CLAUSES (stmt) = ret_clauses; + SET_EXPR_LOCATION (stmt, loc); - tsi_link_before (&i, stmt, TSI_CONTINUE_LINKING); + tsi_link_before (&i, stmt, TSI_CONTINUE_LINKING); + } DECL_ATTRIBUTES (fndecl) = remove_attribute ("oacc declare", DECL_ATTRIBUTES (fndecl)); diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c index 78bcb0a1..41fb35e 100644 --- a/gcc/cp/parser.c +++ b/gcc/cp/parser.c @@ -32123,6 +32123,14 @@ cp_parser_oacc_declare (cp_parser *parser, cp_token *pragma_tok) error = true; continue; } + else if (TREE_PUBLIC (decl)) + { + error_at (loc, + "invalid use of % variable %qD " + "in %<#pragma acc declare%>", decl); + error = true; + continue; + } break; } diff --git a/gcc/testsuite/c-c++-common/goacc/declare-2.c b/gcc/testsuite/c-c++-common/goacc/declare-2.c index ce12463..7979f0c 100644 --- a/gcc/testsuite/c-c++-common/goacc/declare-2.c +++ b/gcc/testsuite/c-c++-common/goacc/declare-2.c @@ -63,4 +63,6 @@ f (void) extern int ve6; #pragma acc declare present_or_create(ve6) /* { dg-error "invalid use of" } */ + +#pragma acc declare present (v9) /* { dg-error "invalid use of" } */ } diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-1.c index 59cfe51..584b921 100644 --- a/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-1.c +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-1.c @@ -4,6 +4,26 @@ #include #include +#define N 8 + +void +subr1 (int *a) +{ + int f[N]; +#pragma acc declare copy (f) + +#pragma acc parallel copy (a[0:N]) + { + int i; + + for (i = 0; i < N; i++) + { + f[i] = a[i]; + a[i] = f[i] + f[i]; + } + } +} + int b[8]; #pragma acc declare create (b) @@ -13,7 +33,6 @@ int d[8] = { 1, 2, 3, 4, 5, 6, 7, 8 }; int main (int argc, char **argv) { - const int N = 8; int a[N]; int e[N]; #pragma acc declare create (e) @@ -61,5 +80,18 @@ main (int argc, char **argv) abort (); } + for (i = 0; i < N; i++) + { + a[i] = 1234; + } + + subr1 (&a[0]); + + for (i = 0; i < N; i++) + { + if (a[i] != 1234 * 2) + abort (); + } + return 0; } diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-5.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-5.c new file mode 100644 index 0000000..8bb7dec --- /dev/null +++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-5.c @@ -0,0 +1,16 @@ +/* { dg-do run { target openacc_nvidia_accel_selected } } */ + +#include +#include +#include + +#define N 8 + +int +main (int argc, char **argv) +{ + int a[N] __attribute__((unused)); +#pragma acc declare present (a) +} + +/* { dg-shouldfail "" } */ --------------060402060609040808080408--