From patchwork Fri May 12 09:36:19 2017 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 8bit X-Patchwork-Submitter: Thomas Schwinge X-Patchwork-Id: 761508 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@bilbo.ozlabs.org Received: from sourceware.org (server1.sourceware.org [209.132.180.131]) (using TLSv1.2 with cipher ECDHE-RSA-AES256-GCM-SHA384 (256/256 bits)) (No client certificate requested) by ozlabs.org (Postfix) with ESMTPS id 3wPPxw4HRtz9s4q for ; Fri, 12 May 2017 19:36:52 +1000 (AEST) Authentication-Results: ozlabs.org; dkim=pass (1024-bit key; unprotected) header.d=gcc.gnu.org header.i=@gcc.gnu.org header.b="HdHSlCQ4"; dkim-atps=neutral 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:subject:in-reply-to:references:date:message-id:mime-version :content-type:content-transfer-encoding; q=dns; s=default; b=mCl DX6mMvqucwuKS9Rw8ET1MgDBQxJaylwffc5JxTqA80fJxCMuFRQgEZ8AEBFkK3u8 HwRB2kQUFpEsJSxu4F6wyu5CwZ7nZLEkfJrX48lmwUirrPu18HF4QEmfky1hYzMi JvM+mY6kEnFP4hJduYhP8mU0b1C6SFfJjjFfst48= DKIM-Signature: v=1; a=rsa-sha1; c=relaxed; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender:from :to:subject:in-reply-to:references:date:message-id:mime-version :content-type:content-transfer-encoding; s=default; bh=Y7mw7+yoJ 0+MAaMRm6yhWK+xEp8=; b=HdHSlCQ4XQEk/RjNdEJnkkuw3rxb5xakFUEpNCSIc mJpA3++opmWKa+1mvp1hCh52fy3FXpCGbKFUsQYXOoH6bKNsey74o8B2rwu4BNeI fJ7XQUwj3EDqa0sxeGM66BJ+XgJ8wtP4PIffH5k+NIlednV9NZ8vgqcLlrr8OjZk Ek= Received: (qmail 91678 invoked by alias); 12 May 2017 09:36:35 -0000 Mailing-List: contact gcc-patches-help@gcc.gnu.org; run by ezmlm Precedence: bulk List-Id: List-Unsubscribe: List-Archive: List-Post: List-Help: Sender: gcc-patches-owner@gcc.gnu.org Delivered-To: mailing list gcc-patches@gcc.gnu.org Received: (qmail 89118 invoked by uid 89); 12 May 2017 09:36:33 -0000 Authentication-Results: sourceware.org; auth=none X-Virus-Found: No X-Spam-SWARE-Status: No, score=-24.4 required=5.0 tests=AWL, BAYES_00, GIT_PATCH_0, GIT_PATCH_1, GIT_PATCH_2, GIT_PATCH_3, RCVD_IN_DNSWL_NONE, SPF_PASS, URIBL_RED autolearn=ham version=3.3.2 spammy= 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; Fri, 12 May 2017 09:36:28 +0000 Received: from nat-ies.mentorg.com ([192.94.31.2] helo=svr-ies-mbx-01.mgc.mentorg.com) by relay1.mentorg.com with esmtp id 1d96zt-0001id-00 from Thomas_Schwinge@mentor.com ; Fri, 12 May 2017 02:36:29 -0700 Received: from hertz.schwinge.homeip.net (137.202.0.87) by svr-ies-mbx-01.mgc.mentorg.com (139.181.222.1) with Microsoft SMTP Server (TLS) id 15.0.1210.3; Fri, 12 May 2017 10:36:25 +0100 From: Thomas Schwinge To: Jakub Jelinek , Subject: Re: OpenACC C front end maintenance: c_parser_oacc_single_int_clause In-Reply-To: <20170510163228.GF1809@tucnak> References: <5632A573.80002@codesourcery.com> <20151030133753.GA478@tucnak.redhat.com> <563381DF.7030005@codesourcery.com> <20151030170555.GE478@tucnak.redhat.com> <5633DF72.4070006@codesourcery.com> <563A49DD.9070107@redhat.com> <563A8255.3060401@codesourcery.com> <87d1bhzpot.fsf@euler.schwinge.homeip.net> <20170510163228.GF1809@tucnak> User-Agent: Notmuch/0.9-101-g81dad07 (http://notmuchmail.org) Emacs/24.5.1 (x86_64-pc-linux-gnu) Date: Fri, 12 May 2017 11:36:19 +0200 Message-ID: <87fuga2z8s.fsf@hertz.schwinge.homeip.net> MIME-Version: 1.0 X-ClientProxiedBy: svr-ies-mbx-01.mgc.mentorg.com (139.181.222.1) To svr-ies-mbx-01.mgc.mentorg.com (139.181.222.1) Hi! On Wed, 10 May 2017 18:32:28 +0200, Jakub Jelinek wrote: > On Tue, May 09, 2017 at 11:27:14PM +0200, Thomas Schwinge wrote: > > OpenACC C front end maintenance: c_parser_oacc_single_int_clause > Ok. Thanks. Committed to trunk in r247960: commit 641fc3aef896abe9687038abfb5fe186c4f350c8 Author: tschwinge Date: Fri May 12 09:33:18 2017 +0000 OpenACC C front end maintenance: c_parser_oacc_single_int_clause gcc/c/ * c-parser.c (c_parser_omp_clause_num_gangs) (c_parser_omp_clause_num_workers) (c_parser_omp_clause_vector_length): Merge functions into... (c_parser_oacc_single_int_clause): ... this new function. Adjust all users. gcc/testsuite/ * c-c++-common/goacc/parallel-dims-1.c: New file. * c-c++-common/goacc/parallel-dims-2.c: Likewise. git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@247960 138bc75d-0d04-0410-961f-82ee72b054a4 --- gcc/c/ChangeLog | 8 + gcc/c/c-parser.c | 176 ++++++--------------- gcc/testsuite/ChangeLog | 3 + gcc/testsuite/c-c++-common/goacc/parallel-dims-1.c | 8 + gcc/testsuite/c-c++-common/goacc/parallel-dims-2.c | 134 ++++++++++++++++ 5 files changed, 202 insertions(+), 127 deletions(-) Grüße Thomas diff --git gcc/c/ChangeLog gcc/c/ChangeLog index 497c9b9..b12f41d 100644 --- gcc/c/ChangeLog +++ gcc/c/ChangeLog @@ -1,3 +1,11 @@ +2017-05-12 Thomas Schwinge + + * c-parser.c (c_parser_omp_clause_num_gangs) + (c_parser_omp_clause_num_workers) + (c_parser_omp_clause_vector_length): Merge functions into... + (c_parser_oacc_single_int_clause): ... this new function. Adjust + all users. + 2017-05-11 Nathan Sidwell * gimple-parser.c: Don't #include tree-dump.h. diff --git gcc/c/c-parser.c gcc/c/c-parser.c index 9398652..90d2d17 100644 --- gcc/c/c-parser.c +++ gcc/c/c-parser.c @@ -11339,51 +11339,6 @@ c_parser_omp_clause_nowait (c_parser *parser ATTRIBUTE_UNUSED, tree list) return c; } -/* OpenACC: - num_gangs ( expression ) */ - -static tree -c_parser_omp_clause_num_gangs (c_parser *parser, tree list) -{ - location_t num_gangs_loc = c_parser_peek_token (parser)->location; - if (c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>")) - { - location_t expr_loc = c_parser_peek_token (parser)->location; - c_expr expr = c_parser_expression (parser); - expr = convert_lvalue_to_rvalue (expr_loc, expr, false, true); - tree c, t = expr.value; - t = c_fully_fold (t, false, NULL); - - c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, "expected %<)%>"); - - if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) - { - c_parser_error (parser, "expected integer expression"); - return list; - } - - /* Attempt to statically determine when the number isn't positive. */ - c = fold_build2_loc (expr_loc, LE_EXPR, boolean_type_node, t, - build_int_cst (TREE_TYPE (t), 0)); - protected_set_expr_location (c, expr_loc); - if (c == boolean_true_node) - { - warning_at (expr_loc, 0, - "% value must be positive"); - t = integer_one_node; - } - - check_no_duplicate_clause (list, OMP_CLAUSE_NUM_GANGS, "num_gangs"); - - c = build_omp_clause (num_gangs_loc, OMP_CLAUSE_NUM_GANGS); - OMP_CLAUSE_NUM_GANGS_EXPR (c) = t; - OMP_CLAUSE_CHAIN (c) = list; - list = c; - } - - return list; -} - /* OpenMP 2.5: num_threads ( expression ) */ @@ -11671,48 +11626,54 @@ c_parser_omp_clause_is_device_ptr (c_parser *parser, tree list) } /* OpenACC: - num_workers ( expression ) */ + num_gangs ( expression ) + num_workers ( expression ) + vector_length ( expression ) */ static tree -c_parser_omp_clause_num_workers (c_parser *parser, tree list) +c_parser_oacc_single_int_clause (c_parser *parser, omp_clause_code code, + tree list) { - location_t num_workers_loc = c_parser_peek_token (parser)->location; - if (c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>")) + location_t loc = c_parser_peek_token (parser)->location; + + if (!c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>")) + return list; + + location_t expr_loc = c_parser_peek_token (parser)->location; + c_expr expr = c_parser_expression (parser); + expr = convert_lvalue_to_rvalue (expr_loc, expr, false, true); + tree c, t = expr.value; + t = c_fully_fold (t, false, NULL); + + c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, "expected %<)%>"); + + if (t == error_mark_node) + return list; + else if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) { - location_t expr_loc = c_parser_peek_token (parser)->location; - c_expr expr = c_parser_expression (parser); - expr = convert_lvalue_to_rvalue (expr_loc, expr, false, true); - tree c, t = expr.value; - t = c_fully_fold (t, false, NULL); + error_at (expr_loc, "%qs expression must be integral", + omp_clause_code_name[code]); + return list; + } - c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, "expected %<)%>"); - - if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) - { - c_parser_error (parser, "expected integer expression"); - return list; - } - - /* Attempt to statically determine when the number isn't positive. */ - c = fold_build2_loc (expr_loc, LE_EXPR, boolean_type_node, t, + /* Attempt to statically determine when the number isn't positive. */ + c = fold_build2_loc (expr_loc, LE_EXPR, boolean_type_node, t, build_int_cst (TREE_TYPE (t), 0)); - protected_set_expr_location (c, expr_loc); - if (c == boolean_true_node) - { - warning_at (expr_loc, 0, - "% value must be positive"); - t = integer_one_node; - } - - check_no_duplicate_clause (list, OMP_CLAUSE_NUM_WORKERS, "num_workers"); - - c = build_omp_clause (num_workers_loc, OMP_CLAUSE_NUM_WORKERS); - OMP_CLAUSE_NUM_WORKERS_EXPR (c) = t; - OMP_CLAUSE_CHAIN (c) = list; - list = c; + protected_set_expr_location (c, expr_loc); + if (c == boolean_true_node) + { + warning_at (expr_loc, 0, + "%qs value must be positive", + omp_clause_code_name[code]); + t = integer_one_node; } - return list; + check_no_duplicate_clause (list, code, omp_clause_code_name[code]); + + c = build_omp_clause (loc, code); + OMP_CLAUSE_OPERAND (c, 0) = t; + OMP_CLAUSE_CHAIN (c) = list; + return c; } /* OpenACC: @@ -12354,51 +12315,6 @@ c_parser_omp_clause_untied (c_parser *parser ATTRIBUTE_UNUSED, tree list) return c; } -/* OpenACC: - vector_length ( expression ) */ - -static tree -c_parser_omp_clause_vector_length (c_parser *parser, tree list) -{ - location_t vector_length_loc = c_parser_peek_token (parser)->location; - if (c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>")) - { - location_t expr_loc = c_parser_peek_token (parser)->location; - c_expr expr = c_parser_expression (parser); - expr = convert_lvalue_to_rvalue (expr_loc, expr, false, true); - tree c, t = expr.value; - t = c_fully_fold (t, false, NULL); - - c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, "expected %<)%>"); - - if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) - { - c_parser_error (parser, "expected integer expression"); - return list; - } - - /* Attempt to statically determine when the number isn't positive. */ - c = fold_build2_loc (expr_loc, LE_EXPR, boolean_type_node, t, - build_int_cst (TREE_TYPE (t), 0)); - protected_set_expr_location (c, expr_loc); - if (c == boolean_true_node) - { - warning_at (expr_loc, 0, - "% value must be positive"); - t = integer_one_node; - } - - check_no_duplicate_clause (list, OMP_CLAUSE_VECTOR_LENGTH, "vector_length"); - - c = build_omp_clause (vector_length_loc, OMP_CLAUSE_VECTOR_LENGTH); - OMP_CLAUSE_VECTOR_LENGTH_EXPR (c) = t; - OMP_CLAUSE_CHAIN (c) = list; - list = c; - } - - return list; -} - /* OpenMP 4.0: inbranch notinbranch */ @@ -13283,11 +13199,15 @@ c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask, c_name = "link"; break; case PRAGMA_OACC_CLAUSE_NUM_GANGS: - clauses = c_parser_omp_clause_num_gangs (parser, clauses); + clauses = c_parser_oacc_single_int_clause (parser, + OMP_CLAUSE_NUM_GANGS, + clauses); c_name = "num_gangs"; break; case PRAGMA_OACC_CLAUSE_NUM_WORKERS: - clauses = c_parser_omp_clause_num_workers (parser, clauses); + clauses = c_parser_oacc_single_int_clause (parser, + OMP_CLAUSE_NUM_WORKERS, + clauses); c_name = "num_workers"; break; case PRAGMA_OACC_CLAUSE_PRESENT: @@ -13341,7 +13261,9 @@ c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask, c_name, clauses); break; case PRAGMA_OACC_CLAUSE_VECTOR_LENGTH: - clauses = c_parser_omp_clause_vector_length (parser, clauses); + clauses = c_parser_oacc_single_int_clause (parser, + OMP_CLAUSE_VECTOR_LENGTH, + clauses); c_name = "vector_length"; break; case PRAGMA_OACC_CLAUSE_WAIT: diff --git gcc/testsuite/ChangeLog gcc/testsuite/ChangeLog index e1e2641..13241be 100644 --- gcc/testsuite/ChangeLog +++ gcc/testsuite/ChangeLog @@ -1,5 +1,8 @@ 2017-05-12 Thomas Schwinge + * c-c++-common/goacc/parallel-dims-1.c: New file. + * c-c++-common/goacc/parallel-dims-2.c: Likewise. + * c-c++-common/goacc/classify-kernels-unparallelized.c: Adjust. * c-c++-common/goacc/classify-kernels.c: Likewise. * c-c++-common/goacc/kernels-counter-vars-function-scope.c: diff --git gcc/testsuite/c-c++-common/goacc/parallel-dims-1.c gcc/testsuite/c-c++-common/goacc/parallel-dims-1.c new file mode 100644 index 0000000..a85d3d3 --- /dev/null +++ gcc/testsuite/c-c++-common/goacc/parallel-dims-1.c @@ -0,0 +1,8 @@ +/* Valid use of OpenACC parallelism dimensions clauses: num_gangs, num_workers, + vector_length. */ + +void f(int i) +{ +#pragma acc parallel num_gangs(i) num_workers(i) vector_length(i) + ; +} diff --git gcc/testsuite/c-c++-common/goacc/parallel-dims-2.c gcc/testsuite/c-c++-common/goacc/parallel-dims-2.c new file mode 100644 index 0000000..30a3d17 --- /dev/null +++ gcc/testsuite/c-c++-common/goacc/parallel-dims-2.c @@ -0,0 +1,134 @@ +/* Invalid use of OpenACC parallelism dimensions clauses: num_gangs, + num_workers, vector_length. */ + +void acc_kernels(int i) +{ +#pragma acc kernels num_gangs(i) /* { dg-error "'num_gangs' is not valid for '#pragma acc kernels'" } */ + ; +#pragma acc kernels num_workers(i) /* { dg-error "'num_workers' is not valid for '#pragma acc kernels'" } */ + ; +#pragma acc kernels vector_length(i) /* { dg-error "'vector_length' is not valid for '#pragma acc kernels'" } */ + ; +} + +void acc_parallel(int i, float f) +{ +#pragma acc parallel num_gangs /* { dg-error "expected '\\(' before end of line" } */ + ; +#pragma acc parallel num_workers /* { dg-error "expected '\\(' before end of line" } */ + ; +#pragma acc parallel vector_length /* { dg-error "expected '\\(' before end of line" } */ + ; + +#pragma acc parallel num_gangs( /* { dg-error "expected (primary-|)expression before end of line" } */ + ; +#pragma acc parallel num_workers( /* { dg-error "expected (primary-|)expression before end of line" } */ + ; +#pragma acc parallel vector_length( /* { dg-error "expected (primary-|)expression before end of line" } */ + ; + +#pragma acc parallel num_gangs() /* { dg-error "expected (primary-|)expression before '\\)' token" } */ + ; +#pragma acc parallel num_workers() /* { dg-error "expected (primary-|)expression before '\\)' token" } */ + ; +#pragma acc parallel vector_length() /* { dg-error "expected (primary-|)expression before '\\)' token" } */ + ; + +#pragma acc parallel num_gangs(1 /* { dg-error "expected '\\)' before end of line" } */ + ; +#pragma acc parallel num_workers(1 /* { dg-error "expected '\\)' before end of line" } */ + ; +#pragma acc parallel vector_length(1 /* { dg-error "expected '\\)' before end of line" } */ + ; + +#pragma acc parallel num_gangs(i /* { dg-error "expected '\\)' before end of line" } */ + ; +#pragma acc parallel num_workers(i /* { dg-error "expected '\\)' before end of line" } */ + ; +#pragma acc parallel vector_length(i /* { dg-error "expected '\\)' before end of line" } */ + ; + +#pragma acc parallel num_gangs(1 i /* { dg-error "expected '\\)' before 'i'" } */ + ; +#pragma acc parallel num_workers(1 i /* { dg-error "expected '\\)' before 'i'" } */ + ; +#pragma acc parallel vector_length(1 i /* { dg-error "expected '\\)' before 'i'" } */ + ; + +#pragma acc parallel num_gangs(1 i) /* { dg-error "expected '\\)' before 'i'" } */ + ; +#pragma acc parallel num_workers(1 i) /* { dg-error "expected '\\)' before 'i'" } */ + ; +#pragma acc parallel vector_length(1 i) /* { dg-error "expected '\\)' before 'i'" } */ + ; + +#pragma acc parallel num_gangs(1, i /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + /* { dg-bogus "expected '\\)' before end of line" "TODO" { xfail c } .-1 } */ + ; +#pragma acc parallel num_workers(1, i /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + /* { dg-bogus "expected '\\)' before end of line" "TODO" { xfail c } .-1 } */ + ; +#pragma acc parallel vector_length(1, i /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + /* { dg-bogus "expected '\\)' before end of line" "TODO" { xfail c } .-1 } */ + ; + +#pragma acc parallel num_gangs(1, i) /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + ; +#pragma acc parallel num_workers(1, i) /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + ; +#pragma acc parallel vector_length(1, i) /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + ; + +#pragma acc parallel num_gangs(num_gangs) /* { dg-error "'num_gangs' (un|was not )declared" } */ + ; +#pragma acc parallel num_workers(num_workers) /* { dg-error "'num_workers' (un|was not )declared" } */ + ; +#pragma acc parallel vector_length(vector_length) /* { dg-error "'vector_length' (un|was not )declared" } */ + ; + +#pragma acc parallel num_gangs(f) /* { dg-error "'num_gangs' expression must be integral" } */ + ; +#pragma acc parallel num_workers(f) /* { dg-error "'num_workers' expression must be integral" } */ + ; +#pragma acc parallel vector_length(f) /* { dg-error "'vector_length' expression must be integral" } */ + ; + +#pragma acc parallel num_gangs((float) 1) /* { dg-error "'num_gangs' expression must be integral" } */ + ; +#pragma acc parallel num_workers((float) 1) /* { dg-error "'num_workers' expression must be integral" } */ + ; +#pragma acc parallel vector_length((float) 1) /* { dg-error "'vector_length' expression must be integral" } */ + ; + +#pragma acc parallel num_gangs(0) /* { dg-warning "'num_gangs' value must be positive" } */ + ; +#pragma acc parallel num_workers(0) /* { dg-warning "'num_workers' value must be positive" } */ + ; +#pragma acc parallel vector_length(0) /* { dg-warning "'vector_length' value must be positive" } */ + ; + +#pragma acc parallel num_gangs((int) -1.2) /* { dg-warning "'num_gangs' value must be positive" } */ + ; +#pragma acc parallel num_workers((int) -1.2) /* { dg-warning "'num_workers' value must be positive" } */ + ; +#pragma acc parallel vector_length((int) -1.2) /* { dg-warning "'vector_length' value must be positive" } */ + ; + +#pragma acc parallel \ + num_gangs(1) /* { dg-error "too many 'num_gangs' clauses" "" { target c } } */ \ + num_workers(1) /* { dg-error "too many 'num_workers' clauses" "" { target c } } */ \ + vector_length(1) /* { dg-error "too many 'vector_length' clauses" "" { target c } } */ \ + num_workers(1) /* { dg-error "too many 'num_workers' clauses" "" { target c++ } } */ \ + vector_length(1) /* { dg-error "too many 'vector_length' clauses" "" { target c++ } } */ \ + num_gangs(1) /* { dg-error "too many 'num_gangs' clauses" "" { target c++ } } */ + ; + +#pragma acc parallel \ + num_gangs(-1) /* { dg-warning "'num_gangs' value must be positive" } */ \ + num_workers() /* { dg-error "expected (primary-|)expression before '\\)' token" } */ \ + vector_length(abc) /* { dg-error "'abc' (un|was not )declared" } */ \ + num_workers(0.5) /* { dg-error "'num_workers' expression must be integral" } */ \ + vector_length(&acc_parallel) /* { dg-error "'vector_length' expression must be integral" } */ \ + num_gangs( /* { dg-error "expected (primary-|)expression before end of line" "TODO" { xfail c } } */ + ; +} Committed to gomp-4_0-branch in r247961: commit 1c87e0644c7e9f84f62c4b1a4f5bca35fba80b36 Author: tschwinge Date: Fri May 12 09:34:37 2017 +0000 OpenACC C front end maintenance: c_parser_oacc_single_int_clause gcc/c/ * c-parser.c (c_parser_omp_clause_num_gangs) (c_parser_omp_clause_num_workers) (c_parser_omp_clause_vector_length): Merge functions into... (c_parser_oacc_single_int_clause): ... this new function. Adjust all users. gcc/testsuite/ * c-c++-common/goacc/parallel-dims-1.c: New file. * c-c++-common/goacc/parallel-dims-2.c: Likewise. trunk r247960 git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/branches/gomp-4_0-branch@247961 138bc75d-0d04-0410-961f-82ee72b054a4 --- gcc/c/ChangeLog.gomp | 8 + gcc/c/c-parser.c | 172 ++++++--------------- gcc/testsuite/ChangeLog.gomp | 3 + gcc/testsuite/c-c++-common/goacc/parallel-dims-1.c | 9 ++ gcc/testsuite/c-c++-common/goacc/parallel-dims-2.c | 134 ++++++++++++++++ 5 files changed, 202 insertions(+), 124 deletions(-) diff --git gcc/c/ChangeLog.gomp gcc/c/ChangeLog.gomp index c3b6798..3efcc8b 100644 --- gcc/c/ChangeLog.gomp +++ gcc/c/ChangeLog.gomp @@ -1,3 +1,11 @@ +2017-05-12 Thomas Schwinge + + * c-parser.c (c_parser_omp_clause_num_gangs) + (c_parser_omp_clause_num_workers) + (c_parser_omp_clause_vector_length): Merge functions into... + (c_parser_oacc_single_int_clause): ... this new function. Adjust + all users. + 2017-05-04 Cesar Philippidis * c-parser.c (c_parser_omp_clause_name): Add support for if_present. diff --git gcc/c/c-parser.c gcc/c/c-parser.c index 957007e..ef61c5f 100644 --- gcc/c/c-parser.c +++ gcc/c/c-parser.c @@ -11217,50 +11217,6 @@ c_parser_omp_clause_nowait (c_parser *parser ATTRIBUTE_UNUSED, tree list) return c; } -/* OpenACC: - num_gangs ( expression ) */ - -static tree -c_parser_omp_clause_num_gangs (c_parser *parser, tree list) -{ - location_t num_gangs_loc = c_parser_peek_token (parser)->location; - if (c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>")) - { - location_t expr_loc = c_parser_peek_token (parser)->location; - tree c, t = c_parser_expression (parser).value; - mark_exp_read (t); - t = c_fully_fold (t, false, NULL); - - c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, "expected %<)%>"); - - if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) - { - c_parser_error (parser, "expected integer expression"); - return list; - } - - /* Attempt to statically determine when the number isn't positive. */ - c = fold_build2_loc (expr_loc, LE_EXPR, boolean_type_node, t, - build_int_cst (TREE_TYPE (t), 0)); - protected_set_expr_location (c, expr_loc); - if (c == boolean_true_node) - { - warning_at (expr_loc, 0, - "% value must be positive"); - t = integer_one_node; - } - - check_no_duplicate_clause (list, OMP_CLAUSE_NUM_GANGS, "num_gangs"); - - c = build_omp_clause (num_gangs_loc, OMP_CLAUSE_NUM_GANGS); - OMP_CLAUSE_NUM_GANGS_EXPR (c) = t; - OMP_CLAUSE_CHAIN (c) = list; - list = c; - } - - return list; -} - /* OpenMP 2.5: num_threads ( expression ) */ @@ -11542,47 +11498,53 @@ c_parser_omp_clause_is_device_ptr (c_parser *parser, tree list) } /* OpenACC: - num_workers ( expression ) */ + num_gangs ( expression ) + num_workers ( expression ) + vector_length ( expression ) */ static tree -c_parser_omp_clause_num_workers (c_parser *parser, tree list) +c_parser_oacc_single_int_clause (c_parser *parser, omp_clause_code code, + tree list) { - location_t num_workers_loc = c_parser_peek_token (parser)->location; - if (c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>")) + location_t loc = c_parser_peek_token (parser)->location; + + if (!c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>")) + return list; + + location_t expr_loc = c_parser_peek_token (parser)->location; + tree c, t = c_parser_expression (parser).value; + mark_exp_read (t); + t = c_fully_fold (t, false, NULL); + + c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, "expected %<)%>"); + + if (t == error_mark_node) + return list; + else if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) { - location_t expr_loc = c_parser_peek_token (parser)->location; - tree c, t = c_parser_expression (parser).value; - mark_exp_read (t); - t = c_fully_fold (t, false, NULL); + error_at (expr_loc, "%qs expression must be integral", + omp_clause_code_name[code]); + return list; + } - c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, "expected %<)%>"); - - if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) - { - c_parser_error (parser, "expected integer expression"); - return list; - } - - /* Attempt to statically determine when the number isn't positive. */ - c = fold_build2_loc (expr_loc, LE_EXPR, boolean_type_node, t, + /* Attempt to statically determine when the number isn't positive. */ + c = fold_build2_loc (expr_loc, LE_EXPR, boolean_type_node, t, build_int_cst (TREE_TYPE (t), 0)); - protected_set_expr_location (c, expr_loc); - if (c == boolean_true_node) - { - warning_at (expr_loc, 0, - "% value must be positive"); - t = integer_one_node; - } - - check_no_duplicate_clause (list, OMP_CLAUSE_NUM_WORKERS, "num_workers"); - - c = build_omp_clause (num_workers_loc, OMP_CLAUSE_NUM_WORKERS); - OMP_CLAUSE_NUM_WORKERS_EXPR (c) = t; - OMP_CLAUSE_CHAIN (c) = list; - list = c; + protected_set_expr_location (c, expr_loc); + if (c == boolean_true_node) + { + warning_at (expr_loc, 0, + "%qs value must be positive", + omp_clause_code_name[code]); + t = integer_one_node; } - return list; + check_no_duplicate_clause (list, code, omp_clause_code_name[code]); + + c = build_omp_clause (loc, code); + OMP_CLAUSE_OPERAND (c, 0) = t; + OMP_CLAUSE_CHAIN (c) = list; + return c; } /* OpenACC: @@ -12342,50 +12304,6 @@ c_parser_omp_clause_untied (c_parser *parser ATTRIBUTE_UNUSED, tree list) return c; } -/* OpenACC: - vector_length ( expression ) */ - -static tree -c_parser_omp_clause_vector_length (c_parser *parser, tree list) -{ - location_t vector_length_loc = c_parser_peek_token (parser)->location; - if (c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>")) - { - location_t expr_loc = c_parser_peek_token (parser)->location; - tree c, t = c_parser_expression (parser).value; - mark_exp_read (t); - t = c_fully_fold (t, false, NULL); - - c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, "expected %<)%>"); - - if (!INTEGRAL_TYPE_P (TREE_TYPE (t))) - { - c_parser_error (parser, "expected integer expression"); - return list; - } - - /* Attempt to statically determine when the number isn't positive. */ - c = fold_build2_loc (expr_loc, LE_EXPR, boolean_type_node, t, - build_int_cst (TREE_TYPE (t), 0)); - protected_set_expr_location (c, expr_loc); - if (c == boolean_true_node) - { - warning_at (expr_loc, 0, - "% value must be positive"); - t = integer_one_node; - } - - check_no_duplicate_clause (list, OMP_CLAUSE_VECTOR_LENGTH, "vector_length"); - - c = build_omp_clause (vector_length_loc, OMP_CLAUSE_VECTOR_LENGTH); - OMP_CLAUSE_VECTOR_LENGTH_EXPR (c) = t; - OMP_CLAUSE_CHAIN (c) = list; - list = c; - } - - return list; -} - /* OpenMP 4.0: inbranch notinbranch */ @@ -13292,11 +13210,15 @@ c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask, c_name = "nohost"; break; case PRAGMA_OACC_CLAUSE_NUM_GANGS: - clauses = c_parser_omp_clause_num_gangs (parser, clauses); + clauses = c_parser_oacc_single_int_clause (parser, + OMP_CLAUSE_NUM_GANGS, + clauses); c_name = "num_gangs"; break; case PRAGMA_OACC_CLAUSE_NUM_WORKERS: - clauses = c_parser_omp_clause_num_workers (parser, clauses); + clauses = c_parser_oacc_single_int_clause (parser, + OMP_CLAUSE_NUM_WORKERS, + clauses); c_name = "num_workers"; break; case PRAGMA_OACC_CLAUSE_PRESENT: @@ -13330,7 +13252,9 @@ c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask, c_name, clauses); break; case PRAGMA_OACC_CLAUSE_VECTOR_LENGTH: - clauses = c_parser_omp_clause_vector_length (parser, clauses); + clauses = c_parser_oacc_single_int_clause (parser, + OMP_CLAUSE_VECTOR_LENGTH, + clauses); c_name = "vector_length"; break; case PRAGMA_OACC_CLAUSE_WAIT: diff --git gcc/testsuite/ChangeLog.gomp gcc/testsuite/ChangeLog.gomp index dadff1a..c24820d 100644 --- gcc/testsuite/ChangeLog.gomp +++ gcc/testsuite/ChangeLog.gomp @@ -1,5 +1,8 @@ 2017-05-12 Thomas Schwinge + * c-c++-common/goacc/parallel-dims-1.c: New file. + * c-c++-common/goacc/parallel-dims-2.c: Likewise. + * c-c++-common/goacc/classify-kernels-unparallelized.c: Adjust. * c-c++-common/goacc/classify-kernels.c: Likewise. * c-c++-common/goacc/kernels-acc-loop-reduction.c: Likewise. diff --git gcc/testsuite/c-c++-common/goacc/parallel-dims-1.c gcc/testsuite/c-c++-common/goacc/parallel-dims-1.c new file mode 100644 index 0000000..9e4cfaa --- /dev/null +++ gcc/testsuite/c-c++-common/goacc/parallel-dims-1.c @@ -0,0 +1,9 @@ +/* Valid use of OpenACC parallelism dimensions clauses: num_gangs, num_workers, + vector_length. */ + +void f(int i) +{ +#pragma acc parallel /* { dg-bogus "region is (gang|worker|vector) partitioned" "" { xfail *-*-* } } */ \ + num_gangs(i) num_workers(i) vector_length(i) + ; +} diff --git gcc/testsuite/c-c++-common/goacc/parallel-dims-2.c gcc/testsuite/c-c++-common/goacc/parallel-dims-2.c new file mode 100644 index 0000000..30a3d17 --- /dev/null +++ gcc/testsuite/c-c++-common/goacc/parallel-dims-2.c @@ -0,0 +1,134 @@ +/* Invalid use of OpenACC parallelism dimensions clauses: num_gangs, + num_workers, vector_length. */ + +void acc_kernels(int i) +{ +#pragma acc kernels num_gangs(i) /* { dg-error "'num_gangs' is not valid for '#pragma acc kernels'" } */ + ; +#pragma acc kernels num_workers(i) /* { dg-error "'num_workers' is not valid for '#pragma acc kernels'" } */ + ; +#pragma acc kernels vector_length(i) /* { dg-error "'vector_length' is not valid for '#pragma acc kernels'" } */ + ; +} + +void acc_parallel(int i, float f) +{ +#pragma acc parallel num_gangs /* { dg-error "expected '\\(' before end of line" } */ + ; +#pragma acc parallel num_workers /* { dg-error "expected '\\(' before end of line" } */ + ; +#pragma acc parallel vector_length /* { dg-error "expected '\\(' before end of line" } */ + ; + +#pragma acc parallel num_gangs( /* { dg-error "expected (primary-|)expression before end of line" } */ + ; +#pragma acc parallel num_workers( /* { dg-error "expected (primary-|)expression before end of line" } */ + ; +#pragma acc parallel vector_length( /* { dg-error "expected (primary-|)expression before end of line" } */ + ; + +#pragma acc parallel num_gangs() /* { dg-error "expected (primary-|)expression before '\\)' token" } */ + ; +#pragma acc parallel num_workers() /* { dg-error "expected (primary-|)expression before '\\)' token" } */ + ; +#pragma acc parallel vector_length() /* { dg-error "expected (primary-|)expression before '\\)' token" } */ + ; + +#pragma acc parallel num_gangs(1 /* { dg-error "expected '\\)' before end of line" } */ + ; +#pragma acc parallel num_workers(1 /* { dg-error "expected '\\)' before end of line" } */ + ; +#pragma acc parallel vector_length(1 /* { dg-error "expected '\\)' before end of line" } */ + ; + +#pragma acc parallel num_gangs(i /* { dg-error "expected '\\)' before end of line" } */ + ; +#pragma acc parallel num_workers(i /* { dg-error "expected '\\)' before end of line" } */ + ; +#pragma acc parallel vector_length(i /* { dg-error "expected '\\)' before end of line" } */ + ; + +#pragma acc parallel num_gangs(1 i /* { dg-error "expected '\\)' before 'i'" } */ + ; +#pragma acc parallel num_workers(1 i /* { dg-error "expected '\\)' before 'i'" } */ + ; +#pragma acc parallel vector_length(1 i /* { dg-error "expected '\\)' before 'i'" } */ + ; + +#pragma acc parallel num_gangs(1 i) /* { dg-error "expected '\\)' before 'i'" } */ + ; +#pragma acc parallel num_workers(1 i) /* { dg-error "expected '\\)' before 'i'" } */ + ; +#pragma acc parallel vector_length(1 i) /* { dg-error "expected '\\)' before 'i'" } */ + ; + +#pragma acc parallel num_gangs(1, i /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + /* { dg-bogus "expected '\\)' before end of line" "TODO" { xfail c } .-1 } */ + ; +#pragma acc parallel num_workers(1, i /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + /* { dg-bogus "expected '\\)' before end of line" "TODO" { xfail c } .-1 } */ + ; +#pragma acc parallel vector_length(1, i /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + /* { dg-bogus "expected '\\)' before end of line" "TODO" { xfail c } .-1 } */ + ; + +#pragma acc parallel num_gangs(1, i) /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + ; +#pragma acc parallel num_workers(1, i) /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + ; +#pragma acc parallel vector_length(1, i) /* { dg-error "expected '\\)' before ',' token" "TODO" { xfail c } } */ + ; + +#pragma acc parallel num_gangs(num_gangs) /* { dg-error "'num_gangs' (un|was not )declared" } */ + ; +#pragma acc parallel num_workers(num_workers) /* { dg-error "'num_workers' (un|was not )declared" } */ + ; +#pragma acc parallel vector_length(vector_length) /* { dg-error "'vector_length' (un|was not )declared" } */ + ; + +#pragma acc parallel num_gangs(f) /* { dg-error "'num_gangs' expression must be integral" } */ + ; +#pragma acc parallel num_workers(f) /* { dg-error "'num_workers' expression must be integral" } */ + ; +#pragma acc parallel vector_length(f) /* { dg-error "'vector_length' expression must be integral" } */ + ; + +#pragma acc parallel num_gangs((float) 1) /* { dg-error "'num_gangs' expression must be integral" } */ + ; +#pragma acc parallel num_workers((float) 1) /* { dg-error "'num_workers' expression must be integral" } */ + ; +#pragma acc parallel vector_length((float) 1) /* { dg-error "'vector_length' expression must be integral" } */ + ; + +#pragma acc parallel num_gangs(0) /* { dg-warning "'num_gangs' value must be positive" } */ + ; +#pragma acc parallel num_workers(0) /* { dg-warning "'num_workers' value must be positive" } */ + ; +#pragma acc parallel vector_length(0) /* { dg-warning "'vector_length' value must be positive" } */ + ; + +#pragma acc parallel num_gangs((int) -1.2) /* { dg-warning "'num_gangs' value must be positive" } */ + ; +#pragma acc parallel num_workers((int) -1.2) /* { dg-warning "'num_workers' value must be positive" } */ + ; +#pragma acc parallel vector_length((int) -1.2) /* { dg-warning "'vector_length' value must be positive" } */ + ; + +#pragma acc parallel \ + num_gangs(1) /* { dg-error "too many 'num_gangs' clauses" "" { target c } } */ \ + num_workers(1) /* { dg-error "too many 'num_workers' clauses" "" { target c } } */ \ + vector_length(1) /* { dg-error "too many 'vector_length' clauses" "" { target c } } */ \ + num_workers(1) /* { dg-error "too many 'num_workers' clauses" "" { target c++ } } */ \ + vector_length(1) /* { dg-error "too many 'vector_length' clauses" "" { target c++ } } */ \ + num_gangs(1) /* { dg-error "too many 'num_gangs' clauses" "" { target c++ } } */ + ; + +#pragma acc parallel \ + num_gangs(-1) /* { dg-warning "'num_gangs' value must be positive" } */ \ + num_workers() /* { dg-error "expected (primary-|)expression before '\\)' token" } */ \ + vector_length(abc) /* { dg-error "'abc' (un|was not )declared" } */ \ + num_workers(0.5) /* { dg-error "'num_workers' expression must be integral" } */ \ + vector_length(&acc_parallel) /* { dg-error "'vector_length' expression must be integral" } */ \ + num_gangs( /* { dg-error "expected (primary-|)expression before end of line" "TODO" { xfail c } } */ + ; +}