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 6C0913851C04; Fri, 18 Feb 2022 20:04:56 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.4.1 sourceware.org 6C0913851C04 Authentication-Results: sourceware.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Authentication-Results: sourceware.org; spf=pass smtp.mailfrom=mentor.com IronPort-SDR: /Hje6ra2KiEzN04pIHu26WtPI5KlEmTxxOvR7s0Y+W9DsWNyM4WQtwWFtC/bIp7Q1Z/DjF2vDo t/58+CmhDRhgdekJ1EBIDDScA+OsvHb/ekP1OByJ2ptK4rn2HVTVJEfCJWuseKbzF5pQ+Qe5+6 yqEEW84WGIgLqnHFZZXR52EQFhDt2C5fkFU3AG6+igcCpelLOu+Lr92g/NuYcVxAL1nVIInQHZ wOuf4tDt2HlNftm74a4R5epqfrQRy+CIDJOrAqr/UOgGIa/iun2kff+z8Sqrh4xSinQ627Y/Ry FrsuKEchY+N4v1kqb87TG5OZ X-IronPort-AV: E=Sophos;i="5.88,379,1635235200"; d="scan'208";a="72177852" Received: from orw-gwy-02-in.mentorg.com ([192.94.38.167]) by esa2.mentor.iphmx.com with ESMTP; 18 Feb 2022 12:04:55 -0800 IronPort-SDR: 61xQPJag9oRssnaOnRIWAr88wzC2HAAI8kqsP0K57kgMs+1eWj1TFmAxxC7FRHEX9tyY8HdOTQ LN6EdTt/ZyslUFqxQtTZhLI3v6aC0Bpvf1tPODqXIB4Y5o71sBLvlqV1zQksB4Q2Ij2D248Pwc eUpc/HoZhaa1yrkKp8y4pUdv4OXEG1te4hpw1jDPFi4DweEr4dq31MhIJNu1Lh8bYwlezI6Scw 0aWrjoc38h4bmipDBAs6jf1Lal1uuOSOnpfhUZPihmmmK4tNBKA3vs2BPwgk6hPyEUXTq1lfjT Xz0= From: Julian Brown To: CC: Jakub Jelinek , Thomas Schwinge , Chung-Lin Tang , Fortran List , Tobias Burnus Subject: [PATCH 5/8] OpenMP: lvalue parsing for map clauses (C++) Date: Fri, 18 Feb 2022 12:04:41 -0800 Message-ID: X-Mailer: git-send-email 2.29.2 In-Reply-To: References: MIME-Version: 1.0 Content-Transfer-Encoding: 8bit Content-Type: text/plain X-Originating-IP: [137.202.0.90] X-ClientProxiedBy: svr-ies-mbx-01.mgc.mentorg.com (139.181.222.1) To SVR-IES-MBX-04.mgc.mentorg.com (139.181.222.4) X-Spam-Status: No, score=-11.8 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, T_SCC_BODY_TEXT_LINE autolearn=ham autolearn_force=no version=3.4.4 X-Spam-Checker-Version: SpamAssassin 3.4.4 (2020-01-24) on server2.sourceware.org X-BeenThere: fortran@gcc.gnu.org X-Mailman-Version: 2.1.29 Precedence: list List-Id: Fortran mailing list List-Unsubscribe: , List-Archive: List-Help: List-Subscribe: , X-List-Received-Date: Fri, 18 Feb 2022 20:05:02 -0000 This patch changes parsing for OpenMP map clauses in C++ to use the generic expression parser, hence adds support for parsing general lvalues (as required by OpenMP 5.0+). So far only a few new types of expression are actually supported throughout compilation (including everything in the testsuite of course, and newly-added tests), and we attempt to reject unsupported expressions in order to avoid surprises for the user. This version of the patch relaxes the check for supported map expressions to include LVALUE_EXPR, which makes pointer-to-member mappings stop "accidentally" failing the check. However such mappings don't actually work yet, so they have been explicitly disallowed elsewhere. Supporting pointer-to-member mappings can probably be done most easily by mapping the whole of the containing struct, and mapping the offsets as firstprivate or similar. That's a little fiddly to get working though, and isn't unambiguously the right thing to do, so that might need more thought. 2022-02-18 Julian Brown gcc/c-family/ * c-omp.cc (c_omp_decompose_attachable_address): Handle more types of expressions. gcc/cp/ * error.cc (dump_expr): Handle OMP_ARRAY_SECTION. * parser.cc (cp_parser_new): Initialize parser->omp_array_section_p. (cp_parser_postfix_open_square_expression): Support OMP_ARRAY_SECTION parsing. (cp_parser_omp_var_list_no_open): Remove ALLOW_DEREF parameter, add MAP_LVALUE in its place. Supported generalised lvalue parsing for map 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. * 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/ * gimplify.cc (accumulate_sibling_list): Handle reference-typed component accesses. Fix support for non-DECL_P struct bases. (omp_ptrmem_p): New function. (omp_build_struct_sibling_lists): Support length-two group for synthesized inner struct mapping. Explicitly disallow pointers to members. * 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/pr67522.C: Likewise. * 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++/ind-base-1.C: New test. * testsuite/libgomp.c++/ind-base-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. --- gcc/c-family/c-omp.cc | 25 ++- gcc/cp/error.cc | 9 + gcc/cp/parser.cc | 141 +++++++++++++-- gcc/cp/parser.h | 3 + gcc/cp/semantics.cc | 55 +++++- gcc/gimplify.cc | 71 +++++++- gcc/testsuite/c-c++-common/gomp/map-6.c | 4 +- gcc/testsuite/g++.dg/gomp/ind-base-3.C | 38 ++++ 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 | 36 ++++ gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C | 39 +++++ .../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 | 86 ++++++++++ gcc/testsuite/g++.dg/gomp/pr67522.C | 2 +- gcc/tree-pretty-print.cc | 14 ++ gcc/tree.def | 3 + libgomp/testsuite/libgomp.c++/ind-base-1.C | 162 ++++++++++++++++++ libgomp/testsuite/libgomp.c++/ind-base-2.C | 49 ++++++ libgomp/testsuite/libgomp.c++/map-comma-1.C | 15 ++ .../testsuite/libgomp.c++/map-rvalue-ref-1.C | 22 +++ .../testsuite/libgomp.c++/member-array-1.C | 89 ++++++++++ libgomp/testsuite/libgomp.c++/struct-ref-1.C | 97 +++++++++++ 25 files changed, 1009 insertions(+), 29 deletions(-) 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++/map-comma-1.C create mode 100644 libgomp/testsuite/libgomp.c++/map-rvalue-ref-1.C create mode 100644 libgomp/testsuite/libgomp.c++/member-array-1.C create mode 100644 libgomp/testsuite/libgomp.c++/struct-ref-1.C diff --git a/gcc/c-family/c-omp.cc b/gcc/c-family/c-omp.cc index e3270f781b0..ad34f817029 100644 --- a/gcc/c-family/c-omp.cc +++ b/gcc/c-family/c-omp.cc @@ -3268,12 +3268,18 @@ c_omp_decompose_attachable_address (tree t, tree *virtbase) { *virtbase = t; - /* It's already a pointer. Just use that. */ - if (POINTER_TYPE_P (TREE_TYPE (t))) + /* It's already a non-offset pointer. Just use that. */ + if (POINTER_TYPE_P (TREE_TYPE (t)) + && (DECL_P (t) + || TREE_CODE (t) == COMPONENT_REF + || TREE_CODE (t) == ARRAY_REF)) return NULL_TREE; /* Otherwise, look for a base pointer deeper within the expression. */ + while (TREE_CODE (t) == COMPOUND_EXPR) + t = TREE_OPERAND (t, 1); + while (TREE_CODE (t) == COMPONENT_REF && (TREE_CODE (TREE_OPERAND (t, 0)) == COMPONENT_REF || TREE_CODE (TREE_OPERAND (t, 0)) == ARRAY_REF)) @@ -3283,9 +3289,24 @@ c_omp_decompose_attachable_address (tree t, tree *virtbase) t = TREE_OPERAND (t, 0); } + if (TREE_CODE (t) == POINTER_PLUS_EXPR) + { + t = TREE_OPERAND (t, 0); + if (TREE_CODE (t) == SAVE_EXPR) + t = TREE_OPERAND (t, 0); + } + + if (TREE_CODE (t) == INDIRECT_REF + && TREE_CODE (TREE_TYPE (TREE_OPERAND (t, 0))) == REFERENCE_TYPE) + t = TREE_OPERAND (t, 0); *virtbase = t; + /* If we have a pointer now (e.g. after we've stripped POINTER_PLUS_EXPR), + we have an offset pointer. That's the attachment point. */ + if (POINTER_TYPE_P (TREE_TYPE (t))) + return t; + if (TREE_CODE (t) != COMPONENT_REF) return NULL_TREE; diff --git a/gcc/cp/error.cc b/gcc/cp/error.cc index e76842e1a2a..a2693aadcd0 100644 --- a/gcc/cp/error.cc +++ b/gcc/cp/error.cc @@ -2436,6 +2436,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 94a5c64be4c..efb65543e11 100644 --- 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. */ + 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; @@ -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; + /* 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, @@ -8038,7 +8045,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)) @@ -8097,6 +8105,32 @@ 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); + + /* 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. */ + 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); 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) + { + 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 the representation previously used for + foo[0]. */ + decl = TREE_OPERAND (decl, 0); + STRIP_NOPS (decl); + + /* ...but don't add the [0:1] representation for references + (because they have special handling elsewhere). */ + 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)) @@ -36622,8 +36739,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) @@ -36709,9 +36825,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++) { @@ -36745,6 +36859,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; @@ -36766,6 +36881,7 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, { *colon = true; cp_parser_require (parser, CPP_COLON, RT_COLON); + finish_scope (); return list; } @@ -36786,6 +36902,7 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind, goto get_comma; } + finish_scope (); return list; } @@ -36794,11 +36911,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; } @@ -36865,7 +36982,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); @@ -40227,12 +40344,12 @@ cp_parser_omp_all_clauses (cp_parser *parser, omp_clause_mask mask, clauses); else clauses = cp_parser_omp_var_list (parser, OMP_CLAUSE_TO, clauses, - true); + false); c_name = "to"; break; case PRAGMA_OMP_CLAUSE_FROM: clauses = cp_parser_omp_var_list (parser, OMP_CLAUSE_FROM, clauses, - true); + false); c_name = "from"; break; case PRAGMA_OMP_CLAUSE_UNIFORM: diff --git a/gcc/cp/parser.h b/gcc/cp/parser.h index d688fd18fd5..7ae507bb135 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 abc0ba3e0ff..a01e79f14b9 100644 --- a/gcc/cp/semantics.cc +++ b/gcc/cp/semantics.cc @@ -5103,7 +5103,9 @@ handle_omp_array_sections_1 (tree c, tree t, vec &types, ret = t_insp.get_deref_toplevel (); 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; @@ -5649,6 +5651,26 @@ handle_omp_array_sections (tree c, enum c_omp_region_type ort) = ((ort != C_ORT_ACC) ? c_omp_decompose_attachable_address (t, &virtbase) : NULL_TREE); + if (TREE_CODE (first) == INDIRECT_REF) + { + /* Detect and skip adding extra nodes for pointer-to-member + mappings. These are unsupported for now. */ + tree tmp = TREE_OPERAND (first, 0); + + if (TREE_CODE (tmp) == NON_LVALUE_EXPR) + tmp = TREE_OPERAND (tmp, 0); + + if (TREE_CODE (tmp) == INDIRECT_REF) + tmp = TREE_OPERAND (tmp, 0); + + if (TREE_CODE (tmp) == POINTER_PLUS_EXPR) + { + tree offset = TREE_OPERAND (tmp, 1); + STRIP_NOPS (offset); + if (TYPE_PTRMEM_P (TREE_TYPE (offset))) + return false; + } + } if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_MAP || (TREE_CODE (t) == COMPONENT_REF && TREE_CODE (TREE_TYPE (t)) == ARRAY_TYPE @@ -5678,7 +5700,9 @@ handle_omp_array_sections (tree c, enum c_omp_region_type ort) bool reference_always_pointer = true; tree c2 = build_omp_clause (OMP_CLAUSE_LOCATION (c), OMP_CLAUSE_MAP); - if (TREE_CODE (t) == COMPONENT_REF) + if (TREE_CODE (t) == COMPONENT_REF + || (TREE_CODE (t) == POINTER_PLUS_EXPR + && TREE_CODE (TREE_OPERAND (t, 0)) == COMPONENT_REF)) { OMP_CLAUSE_SET_MAP_KIND (c2, GOMP_MAP_ATTACH_DETACH); @@ -5708,6 +5732,13 @@ handle_omp_array_sections (tree c, enum c_omp_region_type ort) } OMP_CLAUSE_SET_MAP_KIND (c2, k); } + else if (ort != C_ORT_ACC && attach_pt && !DECL_P (attach_pt)) + { + if (TREE_CODE (TREE_TYPE (t)) != POINTER_TYPE) + return false; + + OMP_CLAUSE_SET_MAP_KIND (c2, GOMP_MAP_ATTACH_DETACH); + } else OMP_CLAUSE_SET_MAP_KIND (c2, GOMP_MAP_FIRSTPRIVATE_POINTER); OMP_CLAUSE_MAP_IMPLICIT (c2) = OMP_CLAUSE_MAP_IMPLICIT (c); @@ -7939,6 +7970,15 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort) { t = t_insp.analyze_components (false); + if (!t_insp.map_supported_p ()) + { + sorry_at (OMP_CLAUSE_LOCATION (c), + "unsupported map expression %qE", + OMP_CLAUSE_DECL (c)); + remove = true; + break; + } + if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP && OMP_CLAUSE_MAP_IMPLICIT (c) && (bitmap_bit_p (&map_head, DECL_UID (t)) @@ -8020,6 +8060,14 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort) OMP_CLAUSE_DECL (c) = t_insp.get_deref_toplevel (); if (type_dependent_expression_p (t_insp.get_deref_toplevel ())) break; + if (!t_insp.map_supported_p ()) + { + sorry_at (OMP_CLAUSE_LOCATION (c), + "unsupported map expression %qE", + OMP_CLAUSE_DECL (c)); + remove = true; + break; + } if (t == error_mark_node) { remove = true; @@ -8049,7 +8097,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/gimplify.cc b/gcc/gimplify.cc index e3b0bd8e8ae..6052add2ab0 100644 --- a/gcc/gimplify.cc +++ b/gcc/gimplify.cc @@ -9823,6 +9823,7 @@ accumulate_sibling_list (enum omp_region_type region_type, enum tree_code code, { tree l = build_omp_clause (OMP_CLAUSE_LOCATION (grp_end), OMP_CLAUSE_MAP); gomp_map_kind k = attach ? GOMP_MAP_FORCE_PRESENT : GOMP_MAP_STRUCT; + tree *tail_chain; OMP_CLAUSE_SET_MAP_KIND (l, k); @@ -9851,9 +9852,13 @@ accumulate_sibling_list (enum omp_region_type region_type, enum tree_code code, { OMP_CLAUSE_CHAIN (extra_node) = *insert_node_pos; OMP_CLAUSE_CHAIN (alloc_node) = extra_node; + tail_chain = &OMP_CLAUSE_CHAIN (extra_node); } else - OMP_CLAUSE_CHAIN (alloc_node) = *insert_node_pos; + { + OMP_CLAUSE_CHAIN (alloc_node) = *insert_node_pos; + tail_chain = &OMP_CLAUSE_CHAIN (alloc_node); + } *insert_node_pos = l; } @@ -9861,6 +9866,7 @@ accumulate_sibling_list (enum omp_region_type region_type, enum tree_code code, { gcc_assert (*grp_start_p == grp_end); grp_start_p = insert_node_after (l, grp_start_p); + tail_chain = &OMP_CLAUSE_CHAIN (*grp_start_p); } tree noind = strip_indirections (base); @@ -9923,8 +9929,7 @@ accumulate_sibling_list (enum omp_region_type region_type, enum tree_code code, while (TREE_CODE (sdecl) == POINTER_PLUS_EXPR) sdecl = TREE_OPERAND (sdecl, 0); - if (DECL_P (sdecl) - && POINTER_TYPE_P (TREE_TYPE (sdecl)) + if (POINTER_TYPE_P (TREE_TYPE (sdecl)) && (region_type & ORT_TARGET)) { tree c2 = build_omp_clause (OMP_CLAUSE_LOCATION (grp_end), @@ -9938,8 +9943,12 @@ accumulate_sibling_list (enum omp_region_type region_type, enum tree_code code, && (TREE_CODE (TREE_TYPE (TREE_OPERAND (TREE_OPERAND (base, 0), 0))) == REFERENCE_TYPE)))); - enum gomp_map_kind mkind = base_ref ? GOMP_MAP_FIRSTPRIVATE_REFERENCE - : GOMP_MAP_FIRSTPRIVATE_POINTER; + enum gomp_map_kind mkind; + if (DECL_P (sdecl)) + mkind = base_ref ? GOMP_MAP_FIRSTPRIVATE_REFERENCE + : GOMP_MAP_FIRSTPRIVATE_POINTER; + else + mkind = GOMP_MAP_ATTACH; OMP_CLAUSE_SET_MAP_KIND (c2, mkind); OMP_CLAUSE_DECL (c2) = sdecl; tree baddr = build_fold_addr_expr (base); @@ -9955,9 +9964,21 @@ accumulate_sibling_list (enum omp_region_type region_type, enum tree_code code, OMP_CLAUSE_SIZE (c2) = fold_build2_loc (OMP_CLAUSE_LOCATION (grp_end), MINUS_EXPR, ptrdiff_type_node, baddr, decladdr); - /* Insert after struct node. */ - OMP_CLAUSE_CHAIN (c2) = OMP_CLAUSE_CHAIN (l); - OMP_CLAUSE_CHAIN (l) = c2; + if (mkind == GOMP_MAP_FIRSTPRIVATE_POINTER + || mkind == GOMP_MAP_FIRSTPRIVATE_REFERENCE) + { + /* Insert after struct node. */ + OMP_CLAUSE_CHAIN (c2) = OMP_CLAUSE_CHAIN (l); + OMP_CLAUSE_CHAIN (l) = c2; + } + else /* GOMP_MAP_ATTACH. */ + { + /* Insert after struct group. */ + OMP_CLAUSE_CHAIN (c2) = *tail_chain; + *tail_chain = c2; + if (*grp_start_p == grp_end) + return &OMP_CLAUSE_CHAIN (*tail_chain); + } } return NULL; @@ -10132,6 +10153,31 @@ accumulate_sibling_list (enum omp_region_type region_type, enum tree_code code, return continue_at; } +/* Return the base (pointer to struct or class) of a pointer-to-member access + expression, or NULL_TREE if EXPR is something else. */ + +static bool +omp_ptrmem_p (tree expr) +{ + if (TREE_CODE (expr) != INDIRECT_REF) + return false; + + expr = TREE_OPERAND (expr, 0); + + if (TREE_CODE (expr) == NON_LVALUE_EXPR) + expr = TREE_OPERAND (expr, 0); + + if (TREE_CODE (expr) == POINTER_PLUS_EXPR) + { + tree base = TREE_OPERAND (expr, 0); + STRIP_NOPS (base); + if (AGGREGATE_TYPE_P (TREE_TYPE (TREE_TYPE (base)))) + return true; + } + + return false; +} + /* Scan through GROUPS, and create sorted structure sibling lists without gimplifying. */ @@ -10194,7 +10240,14 @@ omp_build_struct_sibling_lists (enum tree_code code, STRIP_NOPS (decl); - if (TREE_CODE (decl) != COMPONENT_REF) + if (omp_ptrmem_p (decl)) + { + /* Pointer-to-member mapping types are not yet supported. */ + sorry_at (OMP_CLAUSE_LOCATION (c), "unsupported map expression %qE", + decl); + continue; + } + else if (TREE_CODE (decl) != COMPONENT_REF) continue; omp_mapping_group **wholestruct = NULL; 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/ind-base-3.C b/gcc/testsuite/g++.dg/gomp/ind-base-3.C new file mode 100644 index 00000000000..dbabaf7680c --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/ind-base-3.C @@ -0,0 +1,38 @@ +#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 } */ +/* { dg-message {sorry, unimplemented: unsupported map expression 'choose\(\(& a\), \(& b\), i\)'} "" { target *-*-* } .-2 } */ + 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..cdbd30cdc76 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/map-ptrmem-1.C @@ -0,0 +1,36 @@ +#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: unsupported map expression '\*\(\(\(int\*\*\)\(& s\)\) \+ \(\(sizetype\)ptrp\)\)'} "" { target *-*-* } .-1 } */ + /* { dg-message {sorry, unimplemented: unsupported map expression '\*\(\(\(int\*\)\(& s\)\) \+ \(\(sizetype\)xp\)\)'} "" { target *-*-* } .-2 } */ +#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..b4f735d9f89 --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/map-ptrmem-2.C @@ -0,0 +1,39 @@ +#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: unsupported map expression '\*\(\(\(int\*\*\)s\) \+ \(\(sizetype\)ptrp\)\)'} "" { target *-*-* } .-1 } */ + /* { dg-message {sorry, unimplemented: unsupported map expression '\*\(\(\(int\*\)s\) \+ \(\(sizetype\)xp\)\)'} "" { target *-*-* } .-2 } */ +#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..ba931bd585e --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/member-array-2.C @@ -0,0 +1,86 @@ +#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); + + #pragma omp target map(c.get_arr()[:100]) + #pragma omp teams distribute parallel for + for (int i = 0; i < 100; i++) + c.get_arr()[i] += 2; + + c.check (5, 5); + + #pragma omp target map(c.get_ptr(), c.get_ptr()[:100]) + /* { dg-message {sorry, unimplemented: unsupported map expression 'c\.C::get_ptr\(\)'} "" { target *-*-* } .-1 } */ + #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 666b7a70ea2..86fc5c090d6 100644 --- a/gcc/tree-pretty-print.cc +++ b/gcc/tree-pretty-print.cc @@ -2485,6 +2485,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++/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++/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++/member-array-1.C b/libgomp/testsuite/libgomp.c++/member-array-1.C new file mode 100644 index 00000000000..ee11d45562e --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/member-array-1.C @@ -0,0 +1,89 @@ +/* { dg-do run } */ + +#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); + + #pragma omp target map(c.get_arr()[:100]) + #pragma omp teams distribute parallel for + for (int i = 0; i < 100; i++) + c.get_arr()[i] += 2; + + c.check (5, 5); + + /* This is currently not supported. See also: + gcc/testsuite/g++.dg/gomp/member-array-2.C. */ + //#pragma omp target map(c.get_ptr(), c.get_ptr()[:100]) + //#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/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; +} -- 2.29.2