From mboxrd@z Thu Jan 1 00:00:00 1970 Return-Path: Received: from esa2.mentor.iphmx.com (esa2.mentor.iphmx.com [68.232.141.98]) by sourceware.org (Postfix) with ESMTPS id B433C3858CDA; Tue, 1 Nov 2022 21:50:53 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.4.1 sourceware.org B433C3858CDA Authentication-Results: sourceware.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Authentication-Results: sourceware.org; spf=pass smtp.mailfrom=mentor.com X-IronPort-AV: E=Sophos;i="5.95,232,1661846400"; d="scan'208,223";a="86010742" Received: from orw-gwy-02-in.mentorg.com ([192.94.38.167]) by esa2.mentor.iphmx.com with ESMTP; 01 Nov 2022 13:50:51 -0800 IronPort-SDR: aEVuvQ2fWT/7wQw5K5GXJJ3btm/fDSg3f4WAoiKMCW+/vhsBGOrFumBvINIOIcy4i9dRRJyAt/ 4ER+foTVrv4JCm6e+5cOLp1grNBlTlJbHkUnNT8PCHxUh3jObAiyhE6DvN6xZrc/FQpYB/HwP6 kzszSdwcmlKHcnHlfSvUag8CLs1N4gkgJ6h8caNExgwcr4KbcRutJ8xedGOABbayZduvi+/Jms c8GiE7UyQYrfAOkGsDYjDo3IoTGXTpF/xbNMLdkmBMoojIF6yPYjGxiG4OuDpDeM7UzjyeZtho VP0= Date: Tue, 1 Nov 2022 21:50:38 +0000 From: Julian Brown To: Jakub Jelinek via Fortran CC: Jakub Jelinek , Tobias Burnus , , Thomas Schwinge Subject: Re: [PATCH v2 06/11] OpenMP: lvalue parsing for map clauses (C++) Message-ID: <20221101215038.08a688e1@squid.athome> In-Reply-To: References: <62e4e371468638d2f155c528a5c1e597558a56ac.1647619144.git.julian@codesourcery.com> Organization: Siemens Embedded X-Mailer: Claws Mail 4.1.1git7 (GTK 3.24.34; x86_64-pc-linux-gnu) MIME-Version: 1.0 Content-Type: multipart/mixed; boundary="MP_/Jc+vjnJh=.thHvITP0gzGMK" X-ClientProxiedBy: svr-orw-mbx-11.mgc.mentorg.com (147.34.90.211) To svr-ies-mbx-11.mgc.mentorg.com (139.181.222.11) X-Spam-Status: No, score=-10.5 required=5.0 tests=BAYES_00,GIT_PATCH_0,HEADER_FROM_DIFFERENT_DOMAINS,KAM_DMARC_STATUS,RCVD_IN_MSPIKE_H2,SPF_HELO_PASS,SPF_PASS,TXREP autolearn=ham autolearn_force=no version=3.4.6 X-Spam-Checker-Version: SpamAssassin 3.4.6 (2021-04-09) on server2.sourceware.org List-Id: --MP_/Jc+vjnJh=.thHvITP0gzGMK Content-Type: text/plain; charset="US-ASCII" Content-Transfer-Encoding: 7bit Content-Disposition: inline Hi, On Tue, 24 May 2022 16:15:31 +0200 Jakub Jelinek via Fortran wrote: > On Fri, Mar 18, 2022 at 09:26:47AM -0700, Julian Brown wrote: > > --- a/gcc/cp/parser.cc > > +++ b/gcc/cp/parser.cc > > @@ -4266,6 +4266,9 @@ cp_parser_new (cp_lexer *lexer) > > parser->omp_declare_simd = NULL; > > parser->oacc_routine = NULL; > > > > + /* Allow array slice in expression. */ > > Better /* Disallow OpenMP array sections in expressions. */ Fixed. > > + parser->omp_array_section_p = false; > > + > > /* Not declaring an implicit function template. */ > > parser->auto_is_implicit_function_template_parm_p = false; > > parser->fully_implicit_function_template_p = false; > > I think we should figure out when we should temporarily disable > parser->omp_array_section_p = false; > and restore it afterwards to a saved value. E.g. > cp_parser_lambda_expression seems like a good candidate, the fact that > OpenMP array sections are allowed say in map clause doesn't mean they > are allowed inside of lambdas and it would be especially hard when > the lambda is defining a separate function and the search for > OMP_ARRAY_SECTION probably wouldn't be able to discover those. > Other spots to consider might be statement expressions, perhaps type > definitions etc. I've had a go at doing this -- several expression types now forbid array-section syntax (see new "bad-array-section-*" tests added). I'm afraid my C++ isn't quite up to figuring out how it's possible to define a type inside an expression (inside a map clause) if we forbid lambdas and statement expressions though -- can you give an example? > > @@ -8021,6 +8024,7 @@ cp_parser_postfix_open_square_expression > > (cp_parser *parser, releasing_vec expression_list = NULL; > > location_t loc = cp_lexer_peek_token (parser->lexer)->location; > > bool saved_greater_than_is_operator_p; > > + bool saved_colon_corrects_to_scope_p; > > > > /* Consume the `[' token. */ > > cp_lexer_consume_token (parser->lexer); > > @@ -8028,6 +8032,9 @@ cp_parser_postfix_open_square_expression > > (cp_parser *parser, saved_greater_than_is_operator_p = > > parser->greater_than_is_operator_p; > > parser->greater_than_is_operator_p = true; > > + saved_colon_corrects_to_scope_p = > > parser->colon_corrects_to_scope_p; > > + parser->colon_corrects_to_scope_p = false; > > I think the last above line should be guarded on > if (parser->omp_array_section_p) > There is no reason to get worse diagnostics in non-OpenMP code or > even in OpenMP code where array sections aren't allowed. Fixed. > > + > > + /* NOTE: We are reusing using the type of the whole array as > > the type of > > + the array section here, which isn't necessarily entirely > > correct. > > + Might need revisiting. */ > > "reusing using" looks weird. > As for the type of OMP_ARRAY_SECTION trees, perhaps we could > initially use an incomplete array (so array element would be > meaningful) and when we figure out the details and the array section > is contiguous change its type to array type covering it. This version of the patch makes a best-effort attempt to create an exact-sized array type at parse time, else falls back to an incomplete array type if there are e.g. variable bounds. The type is essentially only used for diagnostics anyway, I think, so that should hopefully be good enough. > > + return build3_loc (input_location, OMP_ARRAY_SECTION, > > + TREE_TYPE (postfix_expression), > > + postfix_expression, index, length); > > + } > > + > > + parser->colon_corrects_to_scope_p = > > saved_colon_corrects_to_scope_p; + > > /* Look for the closing `]'. */ > > cp_parser_require (parser, CPP_CLOSE_SQUARE, RT_CLOSE_SQUARE); > > > > @@ -36536,7 +36570,7 @@ struct omp_dim > > static tree > > cp_parser_omp_var_list_no_open (cp_parser *parser, enum > > omp_clause_code kind, tree list, bool *colon, > > - bool allow_deref = false) > > + bool map_lvalue = false) > > { > > auto_vec dims; > > bool array_section_p; > > @@ -36547,12 +36581,95 @@ cp_parser_omp_var_list_no_open (cp_parser > > *parser, enum omp_clause_code kind, > > parser->colon_corrects_to_scope_p = false; *colon = false; > > } > > + begin_scope (sk_omp, NULL); > > Why? Base-language-wise, clauses don't introduce a new scope > for name-lookup. I think this was in aid of a particular test case (c-c++-common/gomp/map-6.c) that tests various bad usages of "always" and "close" modifiers, together with variables called literally "always" and "close". Parse failures during earlier tests could make later tests fail without the scope. I've moved the scope-creation to the appropriate caller. (Is there a better way? Discarding newly-created symbols on error, perhaps?) > And if it is really needed, I'd strongly prefer to either do it solely > for the clauses that might need it, or do begin_scope before first > such clause and finish at the end if it has been introduced. > > > while (1) > > { > > tree name, decl; > > > > if (kind == OMP_CLAUSE_DEPEND || kind == OMP_CLAUSE_AFFINITY) > > cp_parser_parse_tentatively (parser); > > + else if (map_lvalue && kind == OMP_CLAUSE_MAP) > > + { > > This shouldn't be done just for OMP_CLAUSE_MAP, but for all the > other clauses that accept array sections, including > OMP_CLAUSE_DEPEND, OMP_CLAUSE_AFFINITY, OMP_CLAUSE_MAP, OMP_CLAUSE_TO, > OMP_CLAUSE_FROM, OMP_CLAUSE_INCLUSIVE, OMP_CLAUSE_EXCLUSIVE, > OMP_CLAUSE_USE_DEVICE_ADDR, OMP_CLAUSE_HAS_DEVICE_ADDR, > OMP_CLAUSE_*REDUCTION. I'm not too sure about all of those -- Tobias points out that "INCLUSIVE", "EXCLUSIVE", *DEVICE* and *REDUCTION* take "variable list" item types, not "locator list", though sometimes with an array section being permitted (in OpenMP 5.2+). This version of the patch supports MAP, TO and FROM -- it might be possible to unify the parsing for DEPEND and AFFINITY too, maybe as a later patch. > And preferrably, they should be kept in the IL until > *finish_omp_clauses, which should handle those instead of TREE_LIST > that represented them before. The next patch makes that change. > Additionally, something should diagnose > incorrect uses of OMP_ARRAY_SECTION, which is everywhere in the > expressions but as the outermost node(s), i.e. for clauses that do > allow array sections scan OMP_CLAUSE_DECL after handling handleable > array sections and complain about embedded OMP_ARRAY_SECTION, > including OMP_ARRAY_SECTION say in the lower-bound, length and/or > stride expressions of the valid OMP_ARRAY_SECTION. This version of the patch handles low bound/length incorrectly being array sections, though no extra scan has been needed so far for that (I guess "handle_omp_array_sections" or address parsing in gimplify.cc can handle other cases that might arise). > For C++ that also means handling OMP_ARRAY_SECTION code in pt.c. The next patch handles that bit too. Tested (alongside next patch) with offloading to NVPTX -- with my previously-posted "address tokenization" patch also applied. OK? Thanks, Julian --MP_/Jc+vjnJh=.thHvITP0gzGMK Content-Type: text/x-patch Content-Transfer-Encoding: 7bit Content-Disposition: attachment; filename="0001-OpenMP-lvalue-parsing-for-map-clauses-C.patch" >From eee5df638c2a93759d66dbd15269d849a7616fc4 Mon Sep 17 00:00:00 2001 From: Julian Brown Date: Mon, 12 Sep 2022 17:11:29 +0000 Subject: [PATCH 1/2] OpenMP: lvalue parsing for map clauses (C++) This is a new version of the patch to support "lvalue" parsing (or "locator list item type" parsing), for several OpenMP clause types for C++. All previous review comments have been addressed, hopefully. The patch affects parsing for "map", "to" and "from" clauses: other clause types are handled already (i.e. "affinity" and "depend"), and other clause types mostly accept "variable list" item types, not "locator list" types, so do not need "lvalue" parsing. Several places where array sections should *not* be accepted are now rejected properly at parse time (or semantic analysis time in some cases). 2022-11-01 Julian Brown gcc/c-family/ * c-omp.cc (c_omp_address_inspector::map_supported_p): Handle OMP_ARRAY_SECTION. gcc/cp/ * constexpr.cc (potential_consant_expression_1): Handle OMP_ARRAY_SECTION. * error.cc (dump_expr): Handle OMP_ARRAY_SECTION. * parser.cc (cp_parser_new): Initialize parser->omp_array_section_p. (cp_parser_statement_expr): Disallow array sections. (cp_parser_postfix_open_square_expression): Support OMP_ARRAY_SECTION parsing. (cp_parser_parenthesized_expression_list, cp_parser_lambda_expression, cp_parser_braced_list): Disallow array sections. (cp_parser_omp_var_list_no_open): Remove ALLOW_DEREF parameter, add MAP_LVALUE in its place. Supported generalised lvalue parsing for OpenMP map, to and from clauses. (cp_parser_omp_var_list): Remove ALLOW_DEREF parameter, add MAP_LVALUE. Pass to cp_parser_omp_var_list_no_open. (cp_parser_oacc_data_clause, cp_parser_omp_all_clauses): Update calls to cp_parser_omp_var_list. (cp_parser_omp_clause_map): Add sk_omp scope around cp_parser_omp_var_list_no_open call. * parser.h (cp_parser): Add omp_array_section_p field. * semantics.cc (handle_omp_array_sections_1): Handle more types of map expression. (handle_omp_array_section): Handle non-DECL_P attachment points. (finish_omp_clauses): Check for supported types of expression. gcc/ * tree-pretty-print.c (dump_generic_node): Support OMP_ARRAY_SECTION. * tree.def (OMP_ARRAY_SECTION): New tree code. gcc/testsuite/ * c-c++-common/gomp/map-6.c: Update expected output. * g++.dg/gomp/bad-array-section-1.C: New test. * g++.dg/gomp/bad-array-section-2.C: New test. * g++.dg/gomp/bad-array-section-3.C: New test. * g++.dg/gomp/bad-array-section-4.C: New test. * g++.dg/gomp/bad-array-section-5.C: New test. * g++.dg/gomp/bad-array-section-6.C: New test. * g++.dg/gomp/bad-array-section-7.C: New test. * g++.dg/gomp/bad-array-section-8.C: New test. * g++.dg/gomp/bad-array-section-9.C: New test. * g++.dg/gomp/has_device_addr-non-lvalue-1.C: New test. * g++.dg/gomp/pr67522.C: Update expected output. * g++.dg/gomp/ind-base-3.C: New test. * g++.dg/gomp/map-assignment-1.C: New test. * g++.dg/gomp/map-inc-1.C: New test. * g++.dg/gomp/map-lvalue-ref-1.C: New test. * g++.dg/gomp/map-ptrmem-1.C: New test. * g++.dg/gomp/map-ptrmem-2.C: New test. * g++.dg/gomp/map-static-cast-lvalue-1.C: New test. * g++.dg/gomp/map-ternary-1.C: New test. * g++.dg/gomp/member-array-2.C: New test. libgomp/ * testsuite/libgomp.c++/baseptrs-4.C: Remove commented-out cases that now work. * testsuite/libgomp.c++/ind-base-1.C: New test. * testsuite/libgomp.c++/ind-base-2.C: New test. * testsuite/libgomp.c++/lvalue-tofrom-1.C: New test. * testsuite/libgomp.c++/lvalue-tofrom-2.C: New test. * testsuite/libgomp.c++/map-comma-1.C: New test. * testsuite/libgomp.c++/map-rvalue-ref-1.C: New test. * testsuite/libgomp.c++/member-array-1.C: New test. * testsuite/libgomp.c++/struct-ref-1.C: New test. * testsuite/libgomp.c++/array-field-1.C: New test. * testsuite/libgomp.c++/array-of-struct-1.C: New test. * testsuite/libgomp.c++/array-of-struct-2.C: New test. --- gcc/c-family/c-omp.cc | 1 + gcc/cp/constexpr.cc | 1 + gcc/cp/error.cc | 9 + gcc/cp/parser.cc | 206 +++++++++++++++++- gcc/cp/parser.h | 3 + gcc/cp/semantics.cc | 7 +- gcc/testsuite/c-c++-common/gomp/map-6.c | 4 +- .../g++.dg/gomp/bad-array-section-1.C | 16 ++ .../g++.dg/gomp/bad-array-section-2.C | 16 ++ .../g++.dg/gomp/bad-array-section-3.C | 13 ++ .../g++.dg/gomp/bad-array-section-4.C | 25 +++ .../g++.dg/gomp/bad-array-section-5.C | 29 +++ .../g++.dg/gomp/bad-array-section-6.C | 16 ++ .../g++.dg/gomp/bad-array-section-7.C | 22 ++ .../g++.dg/gomp/bad-array-section-8.C | 26 +++ .../g++.dg/gomp/bad-array-section-9.C | 20 ++ .../gomp/has_device_addr-non-lvalue-1.C | 36 +++ gcc/testsuite/g++.dg/gomp/ind-base-3.C | 37 ++++ gcc/testsuite/g++.dg/gomp/map-assignment-1.C | 12 + gcc/testsuite/g++.dg/gomp/map-inc-1.C | 10 + gcc/testsuite/g++.dg/gomp/map-lvalue-ref-1.C | 19 ++ gcc/testsuite/g++.dg/gomp/map-ptrmem-1.C | 37 ++++ gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C | 40 ++++ .../g++.dg/gomp/map-static-cast-lvalue-1.C | 17 ++ gcc/testsuite/g++.dg/gomp/map-ternary-1.C | 20 ++ gcc/testsuite/g++.dg/gomp/member-array-2.C | 91 ++++++++ gcc/testsuite/g++.dg/gomp/pr67522.C | 2 +- gcc/tree-pretty-print.cc | 14 ++ gcc/tree.def | 3 + libgomp/testsuite/libgomp.c++/baseptrs-4.C | 26 +-- libgomp/testsuite/libgomp.c++/ind-base-1.C | 162 ++++++++++++++ libgomp/testsuite/libgomp.c++/ind-base-2.C | 49 +++++ .../testsuite/libgomp.c++/lvalue-tofrom-1.C | 44 ++++ .../testsuite/libgomp.c++/lvalue-tofrom-2.C | 37 ++++ libgomp/testsuite/libgomp.c++/map-comma-1.C | 15 ++ .../testsuite/libgomp.c++/map-rvalue-ref-1.C | 22 ++ libgomp/testsuite/libgomp.c++/struct-ref-1.C | 97 +++++++++ .../libgomp.c-c++-common/array-field-1.c | 35 +++ .../libgomp.c-c++-common/array-of-struct-1.c | 65 ++++++ .../libgomp.c-c++-common/array-of-struct-2.c | 65 ++++++ 40 files changed, 1338 insertions(+), 31 deletions(-) create mode 100644 gcc/testsuite/g++.dg/gomp/bad-array-section-1.C create mode 100644 gcc/testsuite/g++.dg/gomp/bad-array-section-2.C create mode 100644 gcc/testsuite/g++.dg/gomp/bad-array-section-3.C create mode 100644 gcc/testsuite/g++.dg/gomp/bad-array-section-4.C create mode 100644 gcc/testsuite/g++.dg/gomp/bad-array-section-5.C create mode 100644 gcc/testsuite/g++.dg/gomp/bad-array-section-6.C create mode 100644 gcc/testsuite/g++.dg/gomp/bad-array-section-7.C create mode 100644 gcc/testsuite/g++.dg/gomp/bad-array-section-8.C create mode 100644 gcc/testsuite/g++.dg/gomp/bad-array-section-9.C create mode 100644 gcc/testsuite/g++.dg/gomp/has_device_addr-non-lvalue-1.C create mode 100644 gcc/testsuite/g++.dg/gomp/ind-base-3.C create mode 100644 gcc/testsuite/g++.dg/gomp/map-assignment-1.C create mode 100644 gcc/testsuite/g++.dg/gomp/map-inc-1.C create mode 100644 gcc/testsuite/g++.dg/gomp/map-lvalue-ref-1.C create mode 100644 gcc/testsuite/g++.dg/gomp/map-ptrmem-1.C create mode 100644 gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C create mode 100644 gcc/testsuite/g++.dg/gomp/map-static-cast-lvalue-1.C create mode 100644 gcc/testsuite/g++.dg/gomp/map-ternary-1.C create mode 100644 gcc/testsuite/g++.dg/gomp/member-array-2.C create mode 100644 libgomp/testsuite/libgomp.c++/ind-base-1.C create mode 100644 libgomp/testsuite/libgomp.c++/ind-base-2.C create mode 100644 libgomp/testsuite/libgomp.c++/lvalue-tofrom-1.C create mode 100644 libgomp/testsuite/libgomp.c++/lvalue-tofrom-2.C create mode 100644 libgomp/testsuite/libgomp.c++/map-comma-1.C create mode 100644 libgomp/testsuite/libgomp.c++/map-rvalue-ref-1.C create mode 100644 libgomp/testsuite/libgomp.c++/struct-ref-1.C create mode 100644 libgomp/testsuite/libgomp.c-c++-common/array-field-1.c create mode 100644 libgomp/testsuite/libgomp.c-c++-common/array-of-struct-1.c create mode 100644 libgomp/testsuite/libgomp.c-c++-common/array-of-struct-2.c diff --git a/gcc/c-family/c-omp.cc b/gcc/c-family/c-omp.cc index 7498c883be8..947014e8983 100644 --- a/gcc/c-family/c-omp.cc +++ b/gcc/c-family/c-omp.cc @@ -3228,6 +3228,7 @@ c_omp_address_inspector::map_supported_p () || TREE_CODE (t) == SAVE_EXPR || TREE_CODE (t) == POINTER_PLUS_EXPR || TREE_CODE (t) == NON_LVALUE_EXPR + || TREE_CODE (t) == OMP_ARRAY_SECTION || TREE_CODE (t) == NOP_EXPR) if (TREE_CODE (t) == COMPOUND_EXPR) t = TREE_OPERAND (t, 1); diff --git a/gcc/cp/constexpr.cc b/gcc/cp/constexpr.cc index 03663961bb8..a661096192e 100644 --- a/gcc/cp/constexpr.cc +++ b/gcc/cp/constexpr.cc @@ -9266,6 +9266,7 @@ potential_constant_expression_1 (tree t, bool want_rval, bool strict, bool now, case OACC_ENTER_DATA: case OACC_EXIT_DATA: case OACC_UPDATE: + case OMP_ARRAY_SECTION: /* GCC internal stuff. */ case VA_ARG_EXPR: case TRANSACTION_EXPR: diff --git a/gcc/cp/error.cc b/gcc/cp/error.cc index da8c95c9b43..000f3780ef2 100644 --- a/gcc/cp/error.cc +++ b/gcc/cp/error.cc @@ -2487,6 +2487,15 @@ dump_expr (cxx_pretty_printer *pp, tree t, int flags) pp_cxx_right_bracket (pp); break; + case OMP_ARRAY_SECTION: + dump_expr (pp, TREE_OPERAND (t, 0), flags); + pp_cxx_left_bracket (pp); + dump_expr (pp, TREE_OPERAND (t, 1), flags); + pp_colon (pp); + dump_expr (pp, TREE_OPERAND (t, 2), flags); + pp_cxx_right_bracket (pp); + break; + case UNARY_PLUS_EXPR: dump_unary_op (pp, "+", t, flags); break; diff --git a/gcc/cp/parser.cc b/gcc/cp/parser.cc index a39c5f0d24b..cf619e88f46 100644 --- a/gcc/cp/parser.cc +++ b/gcc/cp/parser.cc @@ -4330,6 +4330,9 @@ cp_parser_new (cp_lexer *lexer) parser->omp_declare_simd = NULL; parser->oacc_routine = NULL; + /* Disallow OpenMP array sections in expressions. */ + parser->omp_array_section_p = false; + /* Not declaring an implicit function template. */ parser->auto_is_implicit_function_template_parm_p = false; parser->fully_implicit_function_template_p = false; @@ -5265,6 +5268,9 @@ static cp_expr cp_parser_statement_expr (cp_parser *parser) { cp_token_position start = cp_parser_start_tentative_firewall (parser); + bool saved_omp_array_section_p = parser->omp_array_section_p; + + parser->omp_array_section_p = false; /* Consume the '('. */ location_t start_loc = cp_lexer_peek_token (parser->lexer)->location; @@ -5281,6 +5287,8 @@ cp_parser_statement_expr (cp_parser *parser) if (!parens.require_close (parser)) cp_parser_skip_to_end_of_statement (parser); + parser->omp_array_section_p = saved_omp_array_section_p; + cp_parser_end_tentative_firewall (parser, start, expr); location_t combined_loc = make_location (start_loc, start_loc, finish_loc); return cp_expr (expr, combined_loc); @@ -8082,6 +8090,7 @@ cp_parser_postfix_open_square_expression (cp_parser *parser, releasing_vec expression_list = NULL; location_t loc = cp_lexer_peek_token (parser->lexer)->location; bool saved_greater_than_is_operator_p; + bool saved_colon_corrects_to_scope_p; /* Consume the `[' token. */ cp_lexer_consume_token (parser->lexer); @@ -8089,6 +8098,10 @@ cp_parser_postfix_open_square_expression (cp_parser *parser, saved_greater_than_is_operator_p = parser->greater_than_is_operator_p; parser->greater_than_is_operator_p = true; + saved_colon_corrects_to_scope_p = parser->colon_corrects_to_scope_p; + if (parser->omp_array_section_p) + parser->colon_corrects_to_scope_p = false; + /* Parse the index expression. */ /* ??? For offsetof, there is a question of what to allow here. If offsetof is not being used in an integral constant expression context, @@ -8099,7 +8112,8 @@ cp_parser_postfix_open_square_expression (cp_parser *parser, constant expressions here. */ if (for_offsetof) index = cp_parser_constant_expression (parser); - else + else if (!parser->omp_array_section_p + || cp_lexer_next_token_is_not (parser->lexer, CPP_COLON)) { if (cxx_dialect >= cxx23 && cp_lexer_next_token_is (parser->lexer, CPP_CLOSE_SQUARE)) @@ -8156,6 +8170,71 @@ cp_parser_postfix_open_square_expression (cp_parser *parser, parser->greater_than_is_operator_p = saved_greater_than_is_operator_p; + if (parser->omp_array_section_p + && cp_lexer_next_token_is (parser->lexer, CPP_COLON)) + { + cp_lexer_consume_token (parser->lexer); + tree length = NULL_TREE; + if (cp_lexer_next_token_is_not (parser->lexer, CPP_CLOSE_SQUARE)) + length = cp_parser_expression (parser); + + parser->colon_corrects_to_scope_p = saved_colon_corrects_to_scope_p; + + if ((index && error_operand_p (index)) + || (length && error_operand_p (length))) + return error_mark_node; + + cp_parser_require (parser, CPP_CLOSE_SQUARE, RT_CLOSE_SQUARE); + + tree idxtype; + + if (index) + index = maybe_constant_value (index); + if (length) + length = maybe_constant_value (length); + + /* If we know the integer bounds, create an index type with exact + low/high (or zero/length) bounds. Otherwise, create an incomplete + array type. (This mostly only affects diagnostics.) */ + if (index != NULL_TREE + && length != NULL_TREE + && TREE_CODE (index) == INTEGER_CST + && TREE_CODE (length) == INTEGER_CST) + { + tree low = fold_convert (sizetype, index); + tree high = fold_convert (sizetype, length); + high = size_binop (PLUS_EXPR, low, high); + high = size_binop (MINUS_EXPR, high, size_one_node); + idxtype = build_range_type (sizetype, low, high); + } + else if ((index == NULL_TREE || integer_zerop (index)) + && length != NULL_TREE + && TREE_CODE (length) == INTEGER_CST) + idxtype = build_index_type (length); + else + idxtype = NULL_TREE; + + tree eltype = ((postfix_expression != error_mark_node + && TREE_TYPE (postfix_expression)) + ? TREE_TYPE (TREE_TYPE (postfix_expression)) + : NULL_TREE); + + tree sectype; + + /* It's not an array or pointer type. Just reuse the type of the + original expression as the type of the array section (an error will be + raised anyway, later). */ + if (eltype == NULL_TREE) + sectype = TREE_TYPE (postfix_expression); + else + sectype = build_array_type (eltype, idxtype); + + return build3_loc (input_location, OMP_ARRAY_SECTION, + sectype, postfix_expression, index, length); + } + + parser->colon_corrects_to_scope_p = saved_colon_corrects_to_scope_p; + /* Look for the closing `]'. */ cp_parser_require (parser, CPP_CLOSE_SQUARE, RT_CLOSE_SQUARE); @@ -8484,6 +8563,7 @@ cp_parser_parenthesized_expression_list (cp_parser* parser, { vec *expression_list; bool saved_greater_than_is_operator_p; + bool saved_omp_array_section_p; /* Assume all the expressions will be constant. */ if (non_constant_p) @@ -8501,6 +8581,9 @@ cp_parser_parenthesized_expression_list (cp_parser* parser, = parser->greater_than_is_operator_p; parser->greater_than_is_operator_p = true; + saved_omp_array_section_p = parser->omp_array_section_p; + parser->omp_array_section_p = false; + cp_expr expr (NULL_TREE); /* Consume expressions until there are no more. */ @@ -8571,6 +8654,7 @@ cp_parser_parenthesized_expression_list (cp_parser* parser, parser->greater_than_is_operator_p = saved_greater_than_is_operator_p; + parser->omp_array_section_p = saved_omp_array_section_p; return expression_list; } @@ -11051,6 +11135,7 @@ cp_parser_lambda_expression (cp_parser* parser) cp_binding_level* implicit_template_scope = parser->implicit_template_scope; bool auto_is_implicit_function_template_parm_p = parser->auto_is_implicit_function_template_parm_p; + bool saved_omp_array_section_p = parser->omp_array_section_p; parser->num_template_parameter_lists = 0; parser->in_statement = 0; @@ -11059,6 +11144,7 @@ cp_parser_lambda_expression (cp_parser* parser) parser->implicit_template_parms = 0; parser->implicit_template_scope = 0; parser->auto_is_implicit_function_template_parm_p = false; + parser->omp_array_section_p = false; /* The body of a lambda in a discarded statement is not discarded. */ bool discarded = in_discarded_stmt; @@ -11111,6 +11197,7 @@ cp_parser_lambda_expression (cp_parser* parser) parser->implicit_template_scope = implicit_template_scope; parser->auto_is_implicit_function_template_parm_p = auto_is_implicit_function_template_parm_p; + parser->omp_array_section_p = saved_omp_array_section_p; } /* This field is only used during parsing of the lambda. */ @@ -25359,6 +25446,9 @@ cp_parser_braced_list (cp_parser* parser, bool* non_constant_p) { tree initializer; location_t start_loc = cp_lexer_peek_token (parser->lexer)->location; + bool saved_omp_array_section_p = parser->omp_array_section_p; + + parser->omp_array_section_p = false; /* Consume the `{' token. */ matching_braces braces; @@ -25391,6 +25481,9 @@ cp_parser_braced_list (cp_parser* parser, bool* non_constant_p) with caret==start at the open brace, finish at the close brace. */ location_t combined_loc = make_location (start_loc, start_loc, finish_loc); result.set_location (combined_loc); + + parser->omp_array_section_p = saved_omp_array_section_p; + return result; } @@ -36760,7 +36853,7 @@ struct omp_dim static tree cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, tree list, bool *colon, - bool allow_deref = false) + bool map_lvalue = false) { auto_vec dims; bool array_section_p; @@ -36777,6 +36870,96 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, if (kind == OMP_CLAUSE_DEPEND || kind == OMP_CLAUSE_AFFINITY) cp_parser_parse_tentatively (parser); + /* This condition doesn't include OMP_CLAUSE_DEPEND or + OMP_CLAUSE_AFFINITY since lvalue ("locator list") parsing for those is + handled further down the function. */ + else if (map_lvalue + && (kind == OMP_CLAUSE_MAP + || kind == OMP_CLAUSE_TO + || kind == OMP_CLAUSE_FROM)) + { + auto s = make_temp_override (parser->omp_array_section_p, true); + token = cp_lexer_peek_token (parser->lexer); + location_t loc = token->location; + decl = cp_parser_assignment_expression (parser); + + dims.truncate (0); + if (TREE_CODE (decl) == OMP_ARRAY_SECTION) + { + while (TREE_CODE (decl) == OMP_ARRAY_SECTION) + { + tree low_bound = TREE_OPERAND (decl, 1); + tree length = TREE_OPERAND (decl, 2); + dims.safe_push (omp_dim (low_bound, length, loc, false)); + decl = TREE_OPERAND (decl, 0); + } + + while (TREE_CODE (decl) == ARRAY_REF + || TREE_CODE (decl) == INDIRECT_REF + || TREE_CODE (decl) == COMPOUND_EXPR) + { + if (REFERENCE_REF_P (decl)) + break; + + if (TREE_CODE (decl) == COMPOUND_EXPR) + { + decl = TREE_OPERAND (decl, 1); + STRIP_NOPS (decl); + } + else if (TREE_CODE (decl) == INDIRECT_REF) + { + dims.safe_push (omp_dim (integer_zero_node, + integer_one_node, loc, true)); + decl = TREE_OPERAND (decl, 0); + } + else /* ARRAY_REF. */ + { + tree index = TREE_OPERAND (decl, 1); + dims.safe_push (omp_dim (index, integer_one_node, loc, + true)); + decl = TREE_OPERAND (decl, 0); + } + } + + /* Bare references have their own special handling, so remove + the explicit dereference added by convert_from_reference. */ + if (REFERENCE_REF_P (decl)) + decl = TREE_OPERAND (decl, 0); + + for (int i = dims.length () - 1; i >= 0; i--) + decl = tree_cons (dims[i].low_bound, dims[i].length, decl); + } + else if (TREE_CODE (decl) == INDIRECT_REF) + { + bool ref_p = REFERENCE_REF_P (decl); + + /* Turn *foo into foo[0:1]. */ + decl = TREE_OPERAND (decl, 0); + STRIP_NOPS (decl); + + /* If we have "*foo" and + - it's an indirection of a reference, "unconvert" it, i.e. + strip the indirection (to just "foo"). + - it's an indirection of a pointer, turn it into + "foo[0:1]". */ + if (!ref_p) + decl = tree_cons (integer_zero_node, integer_one_node, decl); + } + else if (TREE_CODE (decl) == ARRAY_REF) + { + tree idx = TREE_OPERAND (decl, 1); + + decl = TREE_OPERAND (decl, 0); + STRIP_NOPS (decl); + + decl = tree_cons (idx, integer_one_node, decl); + } + else if (TREE_CODE (decl) == NON_LVALUE_EXPR + || CONVERT_EXPR_P (decl)) + decl = TREE_OPERAND (decl, 0); + + goto build_clause; + } token = cp_lexer_peek_token (parser->lexer); if (kind != 0 && cp_parser_is_keyword (token, RID_THIS)) @@ -36855,8 +37038,7 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, case OMP_CLAUSE_TO: start_component_ref: while (cp_lexer_next_token_is (parser->lexer, CPP_DOT) - || (allow_deref - && cp_lexer_next_token_is (parser->lexer, CPP_DEREF))) + || cp_lexer_next_token_is (parser->lexer, CPP_DEREF)) { cpp_ttype ttype = cp_lexer_next_token_is (parser->lexer, CPP_DOT) @@ -36942,9 +37124,7 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, || kind == OMP_CLAUSE_TO) && !array_section_p && (cp_lexer_next_token_is (parser->lexer, CPP_DOT) - || (allow_deref - && cp_lexer_next_token_is (parser->lexer, - CPP_DEREF)))) + || cp_lexer_next_token_is (parser->lexer, CPP_DEREF))) { for (unsigned i = 0; i < dims.length (); i++) { @@ -36978,6 +37158,7 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, cp_parser_parse_definitely (parser); } + build_clause: tree u = build_omp_clause (token->location, kind); OMP_CLAUSE_DECL (u) = decl; OMP_CLAUSE_CHAIN (u) = list; @@ -37027,11 +37208,11 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, static tree cp_parser_omp_var_list (cp_parser *parser, enum omp_clause_code kind, tree list, - bool allow_deref = false) + bool map_lvalue = false) { if (cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN)) return cp_parser_omp_var_list_no_open (parser, kind, list, NULL, - allow_deref); + map_lvalue); return list; } @@ -37098,7 +37279,7 @@ cp_parser_oacc_data_clause (cp_parser *parser, pragma_omp_clause c_kind, gcc_unreachable (); } tree nl, c; - nl = cp_parser_omp_var_list (parser, OMP_CLAUSE_MAP, list, true); + nl = cp_parser_omp_var_list (parser, OMP_CLAUSE_MAP, list, false); for (c = nl; c != list; c = OMP_CLAUSE_CHAIN (c)) OMP_CLAUSE_SET_MAP_KIND (c, kind); @@ -39922,8 +40103,13 @@ cp_parser_omp_clause_map (cp_parser *parser, tree list) cp_lexer_consume_token (parser->lexer); } + /* We introduce a scope here so that errors parsing e.g. "always", "close" + tokens do not propagate to later directives that might use them + legally. */ + begin_scope (sk_omp, NULL); nlist = cp_parser_omp_var_list_no_open (parser, OMP_CLAUSE_MAP, list, NULL, true); + finish_scope (); for (c = nlist; c != list; c = OMP_CLAUSE_CHAIN (c)) OMP_CLAUSE_SET_MAP_KIND (c, kind); diff --git a/gcc/cp/parser.h b/gcc/cp/parser.h index 3e95bfc131b..0104703c09e 100644 --- a/gcc/cp/parser.h +++ b/gcc/cp/parser.h @@ -404,6 +404,9 @@ struct GTY(()) cp_parser { /* TRUE if omp::directive or omp::sequence attributes may not appear. */ bool omp_attrs_forbidden_p; + /* TRUE if an OpenMP array section is allowed. */ + bool omp_array_section_p; + /* Tracks the function's template parameter list when declaring a function using generic type parameters. This is either a new chain in the case of a fully implicit function template or an extension of the function's existing diff --git a/gcc/cp/semantics.cc b/gcc/cp/semantics.cc index 98b9ef460c5..4c6495d1754 100644 --- a/gcc/cp/semantics.cc +++ b/gcc/cp/semantics.cc @@ -5161,7 +5161,9 @@ handle_omp_array_sections_1 (tree c, tree t, vec &types, ret = t_refto; if (TREE_CODE (t) == FIELD_DECL) ret = finish_non_static_data_member (t, NULL_TREE, NULL_TREE); - else if (!VAR_P (t) && TREE_CODE (t) != PARM_DECL) + else if (!VAR_P (t) + && (ort == C_ORT_ACC || !EXPR_P (t)) + && TREE_CODE (t) != PARM_DECL) { if (processing_template_decl && TREE_CODE (t) != OVERLOAD) return NULL_TREE; @@ -8162,7 +8164,8 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort) if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP && (OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_POINTER || OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_ALWAYS_POINTER - || OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_ATTACH_DETACH)) + || OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_ATTACH_DETACH + || (ort != C_ORT_ACC && EXPR_P (t)))) break; if (DECL_P (t)) error_at (OMP_CLAUSE_LOCATION (c), diff --git a/gcc/testsuite/c-c++-common/gomp/map-6.c b/gcc/testsuite/c-c++-common/gomp/map-6.c index 6ee59714847..c749db845b0 100644 --- a/gcc/testsuite/c-c++-common/gomp/map-6.c +++ b/gcc/testsuite/c-c++-common/gomp/map-6.c @@ -20,12 +20,12 @@ foo (void) ; #pragma omp target map (close a) /* { dg-error "'close' undeclared" "" { target c } } */ - /* { dg-error "'close' has not been declared" "" { target c++ } .-1 } */ + /* { dg-error "'close' was not declared in this scope" "" { target c++ } .-1 } */ /* { dg-error "expected '\\)' before 'a'" "" { target *-*-* } .-2 } */ ; #pragma omp target map (always a) /* { dg-error "'always' undeclared" "" { target c } } */ - /* { dg-error "'always' has not been declared" "" { target c++ } .-1 } */ + /* { dg-error "'always' was not declared in this scope" "" { target c++ } .-1 } */ /* { dg-error "expected '\\)' before 'a'" "" { target *-*-* } .-2 } */ ; diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-1.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-1.C new file mode 100644 index 00000000000..738e024d46a --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-1.C @@ -0,0 +1,16 @@ +// { dg-do compile } + +int foo (int *ptr); + +int main() +{ + int arr[20]; + // Reject array section as function argument. +#pragma omp target map(foo(arr[3:5])) +// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 } +// { dg-error {expected '\)' before ':' token} "" { target *-*-* } .-2 } +// { dg-error {expected '\)' before '\]' token} "" { target *-*-* } .-3 } +// { dg-error {expected '#pragma omp' clause before '\]' token} "" { target *-*-* } .-4 } + { } + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-2.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-2.C new file mode 100644 index 00000000000..30b37bc44f4 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-2.C @@ -0,0 +1,16 @@ +// { dg-do compile } +// { dg-additional-options "-std=c++11" } + +int main() +{ + int arr[20]; + // Reject array section in lambda function. +#pragma omp target map([&](const int x) -> int* { return arr[0:x]; } (5)) +// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 } +// { dg-error {invalid conversion from 'int' to 'int\*'} "" { target *-*-* } .-2 } +// { dg-error {expected ';' before ':' token} "" { target *-*-* } .-3 } +// { dg-error {expected primary-expression before ':' token} "" { target *-*-* } .-4 } +// { dg-message {sorry, unimplemented: unsupported map expression 'main\(\)::\{arr\}.main\(\)::\(5\)'} "" { target *-*-* } .-5 } + { } + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-3.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-3.C new file mode 100644 index 00000000000..98dc4a3cf6c --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-3.C @@ -0,0 +1,13 @@ +// { dg-do compile } + +int main() +{ + int arr[20]; + // Reject array section in statement expression. +#pragma omp target map( ({ int x = 5; arr[0:x]; }) ) +// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 } +// { dg-error {expected ';' before ':' token} "" { target *-*-* } .-2 } +// { dg-message {sorry, unimplemented: unsupported map expression '\(\{\.\.\.\}\)'} "" { target *-*-* } .-3 } + { } + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-4.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-4.C new file mode 100644 index 00000000000..ed61fd441d0 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-4.C @@ -0,0 +1,25 @@ +// { dg-do compile } + +struct S { + int *ptr; +}; + +int main() +{ + int arr[20]; + + // Reject array section in compound initialiser. +#pragma omp target map( (struct S) { .ptr = (int *) arr[5:5] } ) +// { dg-error {expected '\]' before ':' token} "" { target *-*-* } .-1 } +// { dg-warning {cast to pointer from integer of different size} "" { target *-*-* } .-2 } +// { dg-error {expected primary-expression before 'struct'} "" { target *-*-* } .-3 } +// { dg-error {expected '\)' before 'struct'} "" { target *-*-* } .-4 } + { } + + // ...and this is unsupported too (probably not useful anyway). +#pragma omp target map( (struct S) { .ptr = &arr[5] } ) +// { dg-message {sorry, unimplemented: unsupported map expression 'S\{\(\& arr\[5\]\)\}'} "" { target *-*-* } .-1 } + { } + + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-5.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-5.C new file mode 100644 index 00000000000..6453351ccde --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-5.C @@ -0,0 +1,29 @@ +// { dg-do compile } + +struct S { + int *ptr; +}; + +int x; + +int main() +{ + int arr[20]; + int *ptr; + /* "arr[1:10]" looks like it might be an expression of array type, hence + able to be indexed (again). This isn't allowed, though. */ +#pragma omp target map(arr[1:10][2]) +// { dg-error {'arr\[1\]' does not have pointer or array type} "" { target *-*-* } .-1 } + { } +#pragma omp target map(arr[1:x][2]) +// { dg-error {'arr\[1\]' does not have pointer or array type} "" { target *-*-* } .-1 } + { } + // ...and nor is this. +#pragma omp target map(ptr[1:10][2]) +// { dg-error {'\*\(ptr \+ [0-9]+\)' does not have pointer or array type} "" { target *-*-* } .-1 } + { } +#pragma omp target map(ptr[1:x][2]) +// { dg-error {'\*\(ptr \+ [0-9]+\)' does not have pointer or array type} "" { target *-*-* } .-1 } + { } + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-6.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-6.C new file mode 100644 index 00000000000..7dee527261d --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-6.C @@ -0,0 +1,16 @@ +// { dg-do compile } + +struct S { + int *ptr; +}; + +bool partly = false; + +int main() +{ + int arr[20]; +#pragma omp target map(partly ? arr[5:5] : arr) +// { dg-message {sorry, unimplemented: unsupported map expression '\(partly \? \(\(int\*\)\(\& arr\[5:5\]\)\) : \(\(int\*\)\(\& arr\)\)\)'} "" { target *-*-* } .-1 } + { } + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-7.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-7.C new file mode 100644 index 00000000000..712a1785338 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-7.C @@ -0,0 +1,22 @@ +// { dg-do compile } + +struct S { + int *ptr; +}; + +int x; + +int main() +{ + int arr[20]; + // Here we know the type of the array section (the upper bound is reported)... +#pragma omp target map(arr[5:5] * 2) +// { dg-error {invalid operands of types 'int \[10\]' and 'int'} "" { target *-*-* } .-1 } + { } + // ...but here, we have an incomplete array type because of the variable + // low bound 'x'. +#pragma omp target map(arr[x:5] * 2) +// { dg-error {invalid operands of types 'int \[\]' and 'int'} "" { target *-*-* } .-1 } + { } + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-8.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-8.C new file mode 100644 index 00000000000..7ee1acb3221 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-8.C @@ -0,0 +1,26 @@ +// { dg-do compile } + +struct T { + int arr[20]; +}; + +struct S { + struct T *tvec; +}; + +int x; + +int main() +{ + struct S *s; + // You can't use an array section like this. Make sure sensible errors are + // reported. +#pragma omp target map(s->tvec[3:5].arr[0:20]) +// { dg-error {request for member 'arr' in 's->S::tvec\[3:5\]', which is of non-class type 'T \[8\]'} "" { target *-*-* } .-1 } + { } +#pragma omp target map(s->tvec[5:x].arr[0:20]) +// { dg-error {invalid use of array with unspecified bounds} "" { target *-*-* } .-1 } +// { dg-error {expected '\)' before 'arr'} "" { target *-*-* } .-2 } + { } + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/bad-array-section-9.C b/gcc/testsuite/g++.dg/gomp/bad-array-section-9.C new file mode 100644 index 00000000000..3198b5f608d --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/bad-array-section-9.C @@ -0,0 +1,20 @@ +// { dg-do compile } + +int x; + +int main() +{ + int arr1[40]; + int arr2[40]; +#pragma omp target map(arr1[arr2[4:5]:arr2[6:7]]) +// { dg-error {low bound 'arr2\[4:5\]' of array section does not have integral type} "" { target *-*-* } .-1 } + { } +#pragma omp target map(arr1[arr2[:1]:arr2[6:1]]) +// { dg-error {low bound 'arr2\[:1\]' of array section does not have integral type} "" { target *-*-* } .-1 } + { } +#pragma omp target map(arr1[x:arr2[6:1]]) +// { dg-error {length 'arr2\[6:1\]' of array section does not have integral type} "" { target *-*-* } .-1 } + { } + return 0; +} + diff --git a/gcc/testsuite/g++.dg/gomp/has_device_addr-non-lvalue-1.C b/gcc/testsuite/g++.dg/gomp/has_device_addr-non-lvalue-1.C new file mode 100644 index 00000000000..3d778538d3a --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/has_device_addr-non-lvalue-1.C @@ -0,0 +1,36 @@ +// { dg-do compile } + +#include +#include +#include + +typedef struct { + int arr[100]; +} S; + +int main() +{ + S *s = new S; + + memset (s->arr, '\0', sizeof s->arr); + +#pragma omp target enter data map(to: (*s).arr) + /* You can't do this, at least as of OpenMP 5.2. "has_device_addr" takes + a "variable list" item type + (OpenMP 5.2, "5.4.9 has_device_addr Clause"). */ +#pragma omp target has_device_addr((*s).arr[5:20]) +// { dg-error {expected unqualified-id before '\(' token} "" { target *-*-* } .-1 } + { + for (int i = 5; i < 25; i++) + s->arr[i] = i; + } + +#pragma omp target exit data map(from: (*s).arr) + + for (int i = 0; i < 100; i++) + assert (i >= 5 && i < 25 ? s->arr[i] == i : s->arr[i] == 0); + + delete s; + + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/ind-base-3.C b/gcc/testsuite/g++.dg/gomp/ind-base-3.C new file mode 100644 index 00000000000..7695b1f907e --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/ind-base-3.C @@ -0,0 +1,37 @@ +#include + +struct S { + int x[10]; +}; + +S * +choose (S *a, S *b, int c) +{ + if (c < 5) + return a; + else + return b; +} + +int main (int argc, char *argv[]) +{ + S a, b; + + for (int i = 0; i < 10; i++) + a.x[i] = b.x[i] = 0; + + for (int i = 0; i < 10; i++) + { +#pragma omp target map(choose(&a, &b, i)->x[:10]) +/* { dg-message {sorry, unimplemented: unsupported map expression 'choose\(\(& a\), \(& b\), i\)->S::x\[0\]'} "" { target *-*-* } .-1 } */ + for (int j = 0; j < 10; j++) + choose (&a, &b, i)->x[j]++; + } + + for (int i = 0; i < 10; i++) + assert (a.x[i] == 5 && b.x[i] == 5); + + return 0; +} + + diff --git a/gcc/testsuite/g++.dg/gomp/map-assignment-1.C b/gcc/testsuite/g++.dg/gomp/map-assignment-1.C new file mode 100644 index 00000000000..5979ec379f1 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/map-assignment-1.C @@ -0,0 +1,12 @@ +#include + +int main (int argc, char *argv[]) +{ + int a = 5, b = 2; +#pragma omp target map(a += b) + /* { dg-message {sorry, unimplemented: unsupported map expression '\(a = \(a \+ b\)\)'} "" { target *-*-* } .-1 } */ + { + a++; + } + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/map-inc-1.C b/gcc/testsuite/g++.dg/gomp/map-inc-1.C new file mode 100644 index 00000000000..b469a4bd548 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/map-inc-1.C @@ -0,0 +1,10 @@ +int main (int argc, char *argv[]) +{ + int a = 5; +#pragma omp target map(++a) + /* { dg-message {sorry, unimplemented: unsupported map expression '\+\+ a'} "" { target *-*-* } .-1 } */ + { + a++; + } + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/map-lvalue-ref-1.C b/gcc/testsuite/g++.dg/gomp/map-lvalue-ref-1.C new file mode 100644 index 00000000000..d720d4318ae --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/map-lvalue-ref-1.C @@ -0,0 +1,19 @@ +#include + +int glob = 10; + +int& foo () +{ + return glob; +} + +int main (int argc, char *argv[]) +{ +#pragma omp target map(foo()) + /* { dg-message {sorry, unimplemented: unsupported map expression 'foo\(\)'} "" { target *-*-* } .-1 } */ + { + foo()++; + } + assert (glob == 11); + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/map-ptrmem-1.C b/gcc/testsuite/g++.dg/gomp/map-ptrmem-1.C new file mode 100644 index 00000000000..c4023f59fc6 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/map-ptrmem-1.C @@ -0,0 +1,37 @@ +#include + +struct S { + int x; + int *ptr; +}; + +int +main (int argc, char *argv[]) +{ + S s; + int S::* xp = &S::x; + int* S::* ptrp = &S::ptr; + + s.ptr = new int[64]; + + s.*xp = 6; + for (int i = 0; i < 64; i++) + (s.*ptrp)[i] = i; + +#pragma omp target map(s.*xp, s.*ptrp, (s.*ptrp)[:64]) + /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\*\(\(\(int\*\*\)\(& s\)\) \+ \(\(sizetype\)ptrp\)\)\)' not supported} "" { target *-*-* } .-1 } */ + /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\(\(int\*\*\)\(& s\)\) \+ \(\(sizetype\)ptrp\)\)' not supported} "" { target *-*-* } .-2 } */ + /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\(\(int\*\)\(& s\)\) \+ \(\(sizetype\)xp\)\)' not supported} "" { target *-*-* } .-3 } */ +#pragma omp teams distribute parallel for + for (int i = 0; i < 64; i++) + { + (s.*xp)++; + (s.*ptrp)[i]++; + } + + assert (s.*xp == 70); + for (int i = 0; i < 64; i++) + assert ((s.*ptrp)[i] == i + 1); + + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C b/gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C new file mode 100644 index 00000000000..fbf379da0eb --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C @@ -0,0 +1,40 @@ +#include + +struct S { + int x; + int *ptr; +}; + +int +main (int argc, char *argv[]) +{ + S *s = new S; + int S::* xp = &S::x; + int* S::* ptrp = &S::ptr; + + s->ptr = new int[64]; + + s->*xp = 4; + for (int i = 0; i < 64; i++) + (s->*ptrp)[i] = i; + +#pragma omp target map(s->*xp, s->*ptrp, (s->*ptrp)[:64]) + /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\(\(int\*\*\)s\) \+ \(\(sizetype\)ptrp\)\)' not supported} "" { target *-*-* } .-1 } */ + /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\(\(int\*\)s\) \+ \(\(sizetype\)xp\)\)' not supported} "" { target *-*-* } .-2 } */ + /* { dg-message {sorry, unimplemented: pointer-to-member mapping '\*\(\*\(\(\(int\*\*\)s\) \+ \(\(sizetype\)ptrp\)\)\)' not supported} "" { target *-*-* } .-3 } */ +#pragma omp teams distribute parallel for + for (int i = 0; i < 64; i++) + { + (s->*xp)++; + (s->*ptrp)[i]++; + } + + assert (s->*xp == 68); + for (int i = 0; i < 64; i++) + assert ((s->*ptrp)[i] == i + 1); + + delete s->ptr; + delete s; + + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/map-static-cast-lvalue-1.C b/gcc/testsuite/g++.dg/gomp/map-static-cast-lvalue-1.C new file mode 100644 index 00000000000..3af9668202c --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/map-static-cast-lvalue-1.C @@ -0,0 +1,17 @@ +#include + +int foo (int x) +{ +#pragma omp target map(static_cast(x)) + /* { dg-message {sorry, unimplemented: unsupported map expression '& x'} "" { target *-*-* } .-1 } */ + { + x += 3; + } + return x; +} + +int main (int argc, char *argv[]) +{ + assert (foo (5) == 8); + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/map-ternary-1.C b/gcc/testsuite/g++.dg/gomp/map-ternary-1.C new file mode 100644 index 00000000000..7b365a909bb --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/map-ternary-1.C @@ -0,0 +1,20 @@ +#include + +int foo (bool yesno) +{ + int x = 5, y = 7; +#pragma omp target map(yesno ? x : y) + /* { dg-message {sorry, unimplemented: unsupported map expression '\(yesno \? x : y\)'} "" { target *-*-* } .-1 } */ + { + x += 3; + y += 5; + } + return yesno ? x : y; +} + +int main (int argc, char *argv[]) +{ + assert (foo (true) == 8); + assert (foo (false) == 12); + return 0; +} diff --git a/gcc/testsuite/g++.dg/gomp/member-array-2.C b/gcc/testsuite/g++.dg/gomp/member-array-2.C new file mode 100644 index 00000000000..caf8ece4262 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/member-array-2.C @@ -0,0 +1,91 @@ +#include + +typedef int intarr100[100]; + +class C { + int arr[100]; + int *ptr; + +public: + C(); + ~C(); + void zero (); + void do_operation (); + void check (int, int); + intarr100 &get_arr () { return arr; } + int *get_ptr() { return ptr; } +}; + +C::C() +{ + ptr = new int[100]; + for (int i = 0; i < 100; i++) + arr[i] = 0; +} + +C::~C() +{ + delete ptr; +} + +void +C::zero () +{ + for (int i = 0; i < 100; i++) + arr[i] = ptr[i] = 0; +} + +void +C::do_operation () +{ +#pragma omp target map(arr, ptr, ptr[:100]) +#pragma omp teams distribute parallel for + for (int i = 0; i < 100; i++) + { + arr[i] = arr[i] + 3; + ptr[i] = ptr[i] + 5; + } +} + +void +C::check (int arrval, int ptrval) +{ + for (int i = 0; i < 100; i++) + { + assert (arr[i] == arrval); + assert (ptr[i] == ptrval); + } +} + +int +main (int argc, char *argv[]) +{ + C c; + + c.zero (); + c.do_operation (); + c.check (3, 5); + + /* It might sort of make sense to be able to do this, but we don't support + it for now. */ + #pragma omp target map(c.get_arr()[:100]) + /* { dg-message {sorry, unimplemented: unsupported map expression 'c\.C::get_arr\(\)\[0\]'} "" { target *-*-* } .-1 } */ + #pragma omp teams distribute parallel for + for (int i = 0; i < 100; i++) + c.get_arr()[i] += 2; + + c.check (5, 5); + + /* Same for this. */ + #pragma omp target map(c.get_ptr(), c.get_ptr()[:100]) + /* { dg-message {sorry, unimplemented: unsupported map expression 'c\.C::get_ptr\(\)'} "" { target *-*-* } .-1 } */ + /* { dg-message {sorry, unimplemented: unsupported map expression '\* c\.C::get_ptr\(\)'} "" { target *-*-* } .-2 } */ + #pragma omp teams distribute parallel for + for (int i = 0; i < 100; i++) + c.get_ptr()[i] += 3; + + c.check (5, 8); + + return 0; +} + diff --git a/gcc/testsuite/g++.dg/gomp/pr67522.C b/gcc/testsuite/g++.dg/gomp/pr67522.C index da8cb74d1fa..4a901ba68c7 100644 --- a/gcc/testsuite/g++.dg/gomp/pr67522.C +++ b/gcc/testsuite/g++.dg/gomp/pr67522.C @@ -12,7 +12,7 @@ foo (void) for (int i = 0; i < 16; i++) ; - #pragma omp target map (S[0:10]) // { dg-error "is not a variable in" } + #pragma omp target map (S[0:10]) // { dg-error "expected primary-expression before '\\\[' token" } ; #pragma omp task depend (inout: S[0:10]) // { dg-error "is not a variable in" } diff --git a/gcc/tree-pretty-print.cc b/gcc/tree-pretty-print.cc index c0656104196..8f95f331779 100644 --- a/gcc/tree-pretty-print.cc +++ b/gcc/tree-pretty-print.cc @@ -2549,6 +2549,20 @@ dump_generic_node (pretty_printer *pp, tree node, int spc, dump_flags_t flags, } break; + case OMP_ARRAY_SECTION: + op0 = TREE_OPERAND (node, 0); + if (op_prio (op0) < op_prio (node)) + pp_left_paren (pp); + dump_generic_node (pp, op0, spc, flags, false); + if (op_prio (op0) < op_prio (node)) + pp_right_paren (pp); + pp_left_bracket (pp); + dump_generic_node (pp, TREE_OPERAND (node, 1), spc, flags, false); + pp_colon (pp); + dump_generic_node (pp, TREE_OPERAND (node, 2), spc, flags, false); + pp_right_bracket (pp); + break; + case CONSTRUCTOR: { unsigned HOST_WIDE_INT ix; diff --git a/gcc/tree.def b/gcc/tree.def index 62650b6934b..f015021e9dc 100644 --- a/gcc/tree.def +++ b/gcc/tree.def @@ -1310,6 +1310,9 @@ DEFTREECODE (OMP_ATOMIC_CAPTURE_NEW, "omp_atomic_capture_new", tcc_statement, 2) /* OpenMP clauses. */ DEFTREECODE (OMP_CLAUSE, "omp_clause", tcc_exceptional, 0) +/* An OpenMP array section. */ +DEFTREECODE (OMP_ARRAY_SECTION, "omp_array_section", tcc_expression, 3) + /* TRANSACTION_EXPR tree code. Operand 0: BODY: contains body of the transaction. */ DEFTREECODE (TRANSACTION_EXPR, "transaction_expr", tcc_expression, 1) diff --git a/libgomp/testsuite/libgomp.c++/baseptrs-4.C b/libgomp/testsuite/libgomp.c++/baseptrs-4.C index 196029ac186..d5ca79c3344 100644 --- a/libgomp/testsuite/libgomp.c++/baseptrs-4.C +++ b/libgomp/testsuite/libgomp.c++/baseptrs-4.C @@ -11,11 +11,9 @@ #define REF2PTR_DECL_BASE #define ARRAY_DECL_BASE -// Needs map clause "lvalue"-parsing support. -//#define REF2ARRAY_DECL_BASE +#define REF2ARRAY_DECL_BASE #define PTR_OFFSET_DECL_BASE -// Needs map clause "lvalue"-parsing support. -//#define REF2PTR_OFFSET_DECL_BASE +#define REF2PTR_OFFSET_DECL_BASE #define MAP_SECTIONS @@ -30,25 +28,21 @@ #define ARRAY_DECL_MEMBER_SLICE #define ARRAY_DECL_MEMBER_SLICE_BASEPTR -// Needs map clause "lvalue"-parsing support. -//#define REF2ARRAY_DECL_MEMBER_SLICE -//#define REF2ARRAY_DECL_MEMBER_SLICE_BASEPTR +#define REF2ARRAY_DECL_MEMBER_SLICE +#define REF2ARRAY_DECL_MEMBER_SLICE_BASEPTR #define PTR_OFFSET_DECL_MEMBER_SLICE #define PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR -// Needs map clause "lvalue"-parsing support. -//#define REF2PTR_OFFSET_DECL_MEMBER_SLICE -//#define REF2PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR +#define REF2PTR_OFFSET_DECL_MEMBER_SLICE +#define REF2PTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR #define PTRARRAY_DECL_MEMBER_SLICE #define PTRARRAY_DECL_MEMBER_SLICE_BASEPTR -// Needs map clause "lvalue"-parsing support. -//#define REF2PTRARRAY_DECL_MEMBER_SLICE -//#define REF2PTRARRAY_DECL_MEMBER_SLICE_BASEPTR +#define REF2PTRARRAY_DECL_MEMBER_SLICE +#define REF2PTRARRAY_DECL_MEMBER_SLICE_BASEPTR #define PTRPTR_OFFSET_DECL_MEMBER_SLICE #define PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR -// Needs map clause "lvalue"-parsing support. -//#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE -//#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR +#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE +#define REF2PTRPTR_OFFSET_DECL_MEMBER_SLICE_BASEPTR #define NONREF_COMPONENT_BASE #define NONREF_COMPONENT_MEMBER_SLICE diff --git a/libgomp/testsuite/libgomp.c++/ind-base-1.C b/libgomp/testsuite/libgomp.c++/ind-base-1.C new file mode 100644 index 00000000000..4566854e60a --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/ind-base-1.C @@ -0,0 +1,162 @@ +// { dg-do run } +// { dg-options "-fopenmp" } + +#include + +struct S +{ + int x[10]; +}; + +struct T +{ + struct S *s; +}; + +struct U +{ + struct T *t; +}; + +void +foo_siblist (void) +{ + U *u = new U; + u->t = new T; + u->t->s = new S; + for (int i = 0; i < 10; i++) + u->t->s->x[i] = 0; +#pragma omp target map(u->t, *(u->t), u->t->s, *u->t->s) + for (int i = 0; i < 10; i++) + u->t->s->x[i] = i * 3; + for (int i = 0; i < 10; i++) + assert (u->t->s->x[i] == i * 3); + delete u->t->s; + delete u->t; + delete u; +} + +void +foo (void) +{ + U *u = new U; + u->t = new T; + u->t->s = new S; + for (int i = 0; i < 10; i++) + u->t->s->x[i] = 0; +#pragma omp target map(*u, u->t, *(u->t), u->t->s, *u->t->s) + for (int i = 0; i < 10; i++) + u->t->s->x[i] = i * 3; + for (int i = 0; i < 10; i++) + assert (u->t->s->x[i] == i * 3); + delete u->t->s; + delete u->t; + delete u; +} + +void +foo_tofrom (void) +{ + U *u = new U; + u->t = new T; + u->t->s = new S; + for (int i = 0; i < 10; i++) + u->t->s->x[i] = 0; +#pragma omp target map(u, *u, u->t, *(u->t), u->t->s, *u->t->s) + for (int i = 0; i < 10; i++) + u->t->s->x[i] = i * 3; + for (int i = 0; i < 10; i++) + assert (u->t->s->x[i] == i * 3); + delete u->t->s; + delete u->t; + delete u; +} + +void +bar (void) +{ + U *u = new U; + U **up = &u; + u->t = new T; + u->t->s = new S; + for (int i = 0; i < 10; i++) + (*up)->t->s->x[i] = 0; +#pragma omp target map(*up, (*up)->t, *(*up)->t, (*up)->t->s, *(*up)->t->s) + for (int i = 0; i < 10; i++) + (*up)->t->s->x[i] = i * 3; + for (int i = 0; i < 10; i++) + assert ((*up)->t->s->x[i] == i * 3); + delete u->t->s; + delete u->t; + delete u; +} + +void +bar_pp (void) +{ + U *u = new U; + U **up = &u; + u->t = new T; + u->t->s = new S; + for (int i = 0; i < 10; i++) + (*up)->t->s->x[i] = 0; +#pragma omp target map(*up, **up, (*up)->t, *(*up)->t, (*up)->t->s, *(*up)->t->s) + for (int i = 0; i < 10; i++) + (*up)->t->s->x[i] = i * 3; + for (int i = 0; i < 10; i++) + assert ((*up)->t->s->x[i] == i * 3); + delete u->t->s; + delete u->t; + delete u; +} + +void +bar_tofrom (void) +{ + U *u = new U; + U **up = &u; + u->t = new T; + u->t->s = new S; + for (int i = 0; i < 10; i++) + (*up)->t->s->x[i] = 0; +#pragma omp target map(*up, up, (*up)->t, *(*up)->t, (*up)->t->s, *(*up)->t->s) + for (int i = 0; i < 10; i++) + (*up)->t->s->x[i] = i * 3; + for (int i = 0; i < 10; i++) + assert ((*up)->t->s->x[i] == i * 3); + delete u->t->s; + delete u->t; + delete u; +} + +void +bar_tofrom_pp (void) +{ + U *u = new U; + U **up = &u; + u->t = new T; + u->t->s = new S; + for (int i = 0; i < 10; i++) + (*up)->t->s->x[i] = 0; +#pragma omp target map(**up, *up, up, (*up)->t, *(*up)->t, (*up)->t->s, \ + *(*up)->t->s) + for (int i = 0; i < 10; i++) + (*up)->t->s->x[i] = i * 3; + for (int i = 0; i < 10; i++) + assert ((*up)->t->s->x[i] == i * 3); + delete u->t->s; + delete u->t; + delete u; +} + +int main (int argc, char *argv[]) +{ + foo_siblist (); + foo (); + foo_tofrom (); + bar (); + bar_pp (); + bar_tofrom (); + bar_tofrom_pp (); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c++/ind-base-2.C b/libgomp/testsuite/libgomp.c++/ind-base-2.C new file mode 100644 index 00000000000..706a1205c00 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/ind-base-2.C @@ -0,0 +1,49 @@ +// { dg-do run } +// { dg-options "-fopenmp" } + +#include + +struct S +{ + int x[10]; +}; + +struct T +{ + struct S ***s; +}; + +struct U +{ + struct T **t; +}; + +void +foo (void) +{ + U *u = new U; + T *real_t = new T; + S *real_s = new S; + T **t_pp = &real_t; + S **s_pp = &real_s; + S ***s_ppp = &s_pp; + u->t = t_pp; + (*u->t)->s = s_ppp; + for (int i = 0; i < 10; i++) + (**((*u->t)->s))->x[i] = 0; +#pragma omp target map(u->t, *u->t, (*u->t)->s, *(*u->t)->s, **(*u->t)->s, \ + (**(*u->t)->s)->x[0:10]) + for (int i = 0; i < 10; i++) + (**((*u->t)->s))->x[i] = i * 3; + for (int i = 0; i < 10; i++) + assert ((**((*u->t)->s))->x[i] == i * 3); + delete real_s; + delete real_t; + delete u; +} + +int main (int argc, char *argv[]) +{ + foo (); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c++/lvalue-tofrom-1.C b/libgomp/testsuite/libgomp.c++/lvalue-tofrom-1.C new file mode 100644 index 00000000000..16534b07bd7 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/lvalue-tofrom-1.C @@ -0,0 +1,44 @@ +#include +#include + +static int lo() +{ + return 30; +} + +static int len() +{ + return 10; +} + +int +main () +{ + char arr[100]; + char *ptr; + + memset (arr, '\0', sizeof arr); + +#pragma omp target enter data map(to: arr[0:100]) + + for (int i = 0; i < 100; i++) + arr[i] = i; + + ptr = &arr[10]; + +#pragma omp target update to(*ptr) + + for (int i = lo (); i < lo () + len (); i++) + arr[i] = i * 2; + +#pragma omp target update to(arr[lo():len()]) + +#pragma omp target exit data map(from: arr[0:100]) + + assert (arr[10] == 10); + for (int i = lo (); i < lo () + len (); i++) + assert (arr[i] == i * 2); + + return 0; +} + diff --git a/libgomp/testsuite/libgomp.c++/lvalue-tofrom-2.C b/libgomp/testsuite/libgomp.c++/lvalue-tofrom-2.C new file mode 100644 index 00000000000..214ca299f62 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/lvalue-tofrom-2.C @@ -0,0 +1,37 @@ +#include +#include +#include + +struct array_wrapper { + int *data; + unsigned int length; +}; + +int +main () +{ + struct array_wrapper aw; + + aw.data = new int[100]; + aw.length = 100; + +#pragma omp target enter data map(to: aw.data, aw.length) \ + map(alloc: aw.data[0:aw.length]) + +#pragma omp target + for (int i = 0; i < aw.length; i++) + aw.data[i] = i; + +#pragma omp target update from(aw.data[:aw.length]) + +#pragma omp target exit data map(delete: aw.data, aw.length, \ + aw.data[0:aw.length]) + + for (int i = 0; i < aw.length; i++) + assert (aw.data[i] == i); + + delete[] aw.data; + + return 0; +} + diff --git a/libgomp/testsuite/libgomp.c++/map-comma-1.C b/libgomp/testsuite/libgomp.c++/map-comma-1.C new file mode 100644 index 00000000000..ee03c5ac1aa --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/map-comma-1.C @@ -0,0 +1,15 @@ +/* { dg-do run } */ + +#include + +int main (int argc, char *argv[]) +{ + int a = 5, b = 7; +#pragma omp target map((a, b)) + { + a++; + b++; + } + assert (a == 5 && b == 8); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c++/map-rvalue-ref-1.C b/libgomp/testsuite/libgomp.c++/map-rvalue-ref-1.C new file mode 100644 index 00000000000..93811da4000 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/map-rvalue-ref-1.C @@ -0,0 +1,22 @@ +/* { dg-do run } */ + +#include + +int foo (int &&x) +{ + int y; +#pragma omp target map(x, y) + { + x++; + y = x; + } + return y; +} + +int main (int argc, char *argv[]) +{ + int y = 5; + y = foo (y + 3); + assert (y == 9); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c++/struct-ref-1.C b/libgomp/testsuite/libgomp.c++/struct-ref-1.C new file mode 100644 index 00000000000..d3874650017 --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/struct-ref-1.C @@ -0,0 +1,97 @@ +// { dg-do run } +// { dg-options "-fopenmp" } + +#include + +struct S +{ + int x[10]; +}; + +void +foo (S *s, int x) +{ + S *&r = s; + for (int i = 0; i < x; i++) + s[0].x[i] = s[1].x[i] = 0; + #pragma omp target map (s, x) + ; + #pragma omp target map (s[0], x) + for (int i = 0; i < x; i++) + s[0].x[i] = i; + #pragma omp target map (s[1], x) + for (int i = 0; i < x; i++) + s[1].x[i] = i * 2; + for (int i = 0; i < x; i++) + { + assert (s[0].x[i] == i); + assert (s[1].x[i] == i * 2); + s[0].x[i] = 0; + s[1].x[i] = 0; + } + #pragma omp target map (r, x) + ; + #pragma omp target map (r[0], x) + for (int i = 0; i < x; i++) + r[0].x[i] = i; + #pragma omp target map (r[1], x) + for (int i = 0; i < x; i++) + r[1].x[i] = i * 2; + for (int i = 0; i < x; i++) + { + assert (r[0].x[i] == i); + assert (r[1].x[i] == i * 2); + } +} + +template +struct T +{ + int x[N]; +}; + +template +void +bar (T *t, int x) +{ + T *&r = t; + for (int i = 0; i < x; i++) + t[0].x[i] = t[1].x[i] = 0; + #pragma omp target map (t, x) + ; + #pragma omp target map (t[0], x) + for (int i = 0; i < x; i++) + t[0].x[i] = i; + #pragma omp target map (t[1], x) + for (int i = 0; i < x; i++) + t[1].x[i] = i * 2; + for (int i = 0; i < x; i++) + { + assert (t[0].x[i] == i); + assert (t[1].x[i] == i * 2); + t[0].x[i] = 0; + t[1].x[i] = 0; + } + #pragma omp target map (r, x) + ; + #pragma omp target map (r[0], x) + for (int i = 0; i < x; i++) + r[0].x[i] = i; + #pragma omp target map (r[1], x) + for (int i = 0; i < x; i++) + r[1].x[i] = i * 2; + for (int i = 0; i < x; i++) + { + assert (r[0].x[i] == i); + assert (r[1].x[i] == i * 2); + } +} + +int main (int argc, char *argv[]) +{ + S s[2]; + foo (s, 10); + T<10> t[2]; + bar (t, 10); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c-c++-common/array-field-1.c b/libgomp/testsuite/libgomp.c-c++-common/array-field-1.c new file mode 100644 index 00000000000..6dd8b5c48e1 --- /dev/null +++ b/libgomp/testsuite/libgomp.c-c++-common/array-field-1.c @@ -0,0 +1,35 @@ +/* { dg-do run } */ + +#include +#include +#include + +#define N 16 + +struct Z { + int *ptr; + int arr[N]; + int c; +}; + +int main (int argc, char *argv[]) +{ + struct Z *myz; + myz = (struct Z *) calloc (1, sizeof *myz); + +#pragma omp target map(tofrom:myz->arr[0:N], myz->c) + { + for (int i = 0; i < N; i++) + myz->arr[i]++; + myz->c++; + } + + for (int i = 0; i < N; i++) + assert (myz->arr[i] == 1); + assert (myz->c == 1); + + free (myz); + + return 0; +} + diff --git a/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-1.c b/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-1.c new file mode 100644 index 00000000000..11215b1df7a --- /dev/null +++ b/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-1.c @@ -0,0 +1,65 @@ +/* { dg-do run } */ + +#include +#include +#include + +#define N 16 + +/* NOTE: This test is the same as array-of-struct-2.c, except the fields of + this struct are in a different order. */ + +struct Z { + int arr[N]; + int *ptr; + int c; +}; + +void +foo (struct Z *zarr, int len) +{ +#pragma omp target map(to:zarr, zarr[5].ptr) map(tofrom:zarr[5].ptr[0:len]) + { + for (int i = 0; i < len; i++) + zarr[5].ptr[i]++; + } + +#pragma omp target map(to:zarr) map(tofrom:zarr[4].arr[0:len]) + { + for (int i = 0; i < len; i++) + zarr[4].arr[i]++; + } + +#pragma omp target map (to:zarr[3].ptr) map(tofrom:zarr[3].ptr[0:len]) + { + for (int i = 0; i < len; i++) + zarr[3].ptr[i]++; + } + +#pragma omp target map(tofrom:zarr[2].arr[0:len]) + { + for (int i = 0; i < len; i++) + zarr[2].arr[i]++; + } +} + +int main (int argc, char *argv[]) +{ + struct Z zs[10]; + memset (zs, 0, sizeof zs); + + for (int i = 0; i < 10; i++) + zs[i].ptr = (int *) calloc (N, sizeof (int)); + + foo (zs, N); + + for (int i = 0; i < N; i++) + { + assert (zs[2].arr[i] == 1); + assert (zs[4].arr[i] == 1); + assert (zs[3].ptr[i] == 1); + assert (zs[5].ptr[i] == 1); + } + + return 0; +} diff --git a/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-2.c b/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-2.c new file mode 100644 index 00000000000..d5d74b8c07d --- /dev/null +++ b/libgomp/testsuite/libgomp.c-c++-common/array-of-struct-2.c @@ -0,0 +1,65 @@ +/* { dg-do run } */ + +#include +#include +#include + +#define N 16 + +/* NOTE: This test is the same as array-of-struct-1.c, except the fields of + this struct are in a different order. */ + +struct Z { + int *ptr; + int arr[N]; + int c; +}; + +void +foo (struct Z *zarr, int len) +{ +#pragma omp target map(to:zarr, zarr[5].ptr) map(tofrom:zarr[5].ptr[0:len]) + { + for (int i = 0; i < len; i++) + zarr[5].ptr[i]++; + } + +#pragma omp target map(to:zarr) map(tofrom:zarr[4].arr[0:len]) + { + for (int i = 0; i < len; i++) + zarr[4].arr[i]++; + } + +#pragma omp target map (to:zarr[3].ptr) map(tofrom:zarr[3].ptr[0:len]) + { + for (int i = 0; i < len; i++) + zarr[3].ptr[i]++; + } + +#pragma omp target map(tofrom:zarr[2].arr[0:len]) + { + for (int i = 0; i < len; i++) + zarr[2].arr[i]++; + } +} + +int main (int argc, char *argv[]) +{ + struct Z zs[10]; + memset (zs, 0, sizeof zs); + + for (int i = 0; i < 10; i++) + zs[i].ptr = (int *) calloc (N, sizeof (int)); + + foo (zs, N); + + for (int i = 0; i < N; i++) + { + assert (zs[2].arr[i] == 1); + assert (zs[4].arr[i] == 1); + assert (zs[3].ptr[i] == 1); + assert (zs[5].ptr[i] == 1); + } + + return 0; +} -- 2.29.2 --MP_/Jc+vjnJh=.thHvITP0gzGMK--