aboutsummaryrefslogtreecommitdiff
path: root/gcc/cp/parser.c
diff options
context:
space:
mode:
authorChung-Lin Tang <cltang@codesourcery.com>2021-12-08 22:28:03 +0800
committerChung-Lin Tang <cltang@codesourcery.com>2021-12-08 22:29:06 +0800
commit0ab29cf0bb68960c1f87405f14b4fb2109254e2f (patch)
treedc3a692a82f7a3f39c6b7b1f4dbb00c7088a8fa7 /gcc/cp/parser.c
parentdbf8bd3c2f2cd2d27ca4f0fe379bd9490273c6d7 (diff)
downloadgcc-0ab29cf0bb68960c1f87405f14b4fb2109254e2f.zip
gcc-0ab29cf0bb68960c1f87405f14b4fb2109254e2f.tar.gz
gcc-0ab29cf0bb68960c1f87405f14b4fb2109254e2f.tar.bz2
openmp: Improve OpenMP target support for C++ (PR92120)
This patch implements several C++ specific mapping capabilities introduced for OpenMP 5.0, including implicit mapping of this[:1] for non-static member functions, zero-length array section mapping of pointer-typed members, lambda captured variable access in target regions, and use of lambda objects inside target regions. Several adjustments to the C/C++ front-ends to allow more member-access syntax as valid is also included. PR middle-end/92120 gcc/cp/ChangeLog: * cp-tree.h (finish_omp_target): New declaration. (finish_omp_target_clauses): Likewise. * parser.c (cp_parser_omp_clause_map): Adjust call to cp_parser_omp_var_list_no_open to set 'allow_deref' argument to true. (cp_parser_omp_target): Factor out code, adjust into calls to new function finish_omp_target. * pt.c (tsubst_expr): Add call to finish_omp_target_clauses for OMP_TARGET case. * semantics.c (handle_omp_array_sections_1): Add handling to create 'this->member' from 'member' FIELD_DECL. Remove case of rejecting 'this' when not in declare simd. (handle_omp_array_sections): Likewise. (finish_omp_clauses): Likewise. Adjust to allow 'this[]' in OpenMP map clauses. Handle 'A->member' case in map clauses. Remove case of rejecting 'this' when not in declare simd. (struct omp_target_walk_data): New struct for walking over target-directive tree body. (finish_omp_target_clauses_r): New function for tree walk. (finish_omp_target_clauses): New function. (finish_omp_target): New function. gcc/c/ChangeLog: * c-parser.c (c_parser_omp_clause_map): Set 'allow_deref' argument in call to c_parser_omp_variable_list to 'true'. * c-typeck.c (handle_omp_array_sections_1): Add strip of MEM_REF in array base handling. (c_finish_omp_clauses): Handle 'A->member' case in map clauses. gcc/ChangeLog: * gimplify.c ("tree-hash-traits.h"): Add include. (gimplify_scan_omp_clauses): Change struct_map_to_clause to type hash_map<tree_operand, tree> *. Adjust struct map handling to handle cases of *A and A->B expressions. Under !DECL_P case of GOMP_CLAUSE_MAP handling, add STRIP_NOPS for indir_p case, add to struct_deref_set for map(*ptr_to_struct) cases. Add MEM_REF case when handling component_ref_p case. Add unshare_expr and gimplification when created GOMP_MAP_STRUCT is not a DECL. Add code to add firstprivate pointer for *pointer-to-struct case. (gimplify_adjust_omp_clauses): Move GOMP_MAP_STRUCT removal code for exit data directives code to earlier position. * omp-low.c (lower_omp_target): Handle GOMP_MAP_ATTACH_ZERO_LENGTH_ARRAY_SECTION, and GOMP_MAP_POINTER_TO_ZERO_LENGTH_ARRAY_SECTION map kinds. * tree-pretty-print.c (dump_omp_clause): Likewise. gcc/testsuite/ChangeLog: * gcc.dg/gomp/target-3.c: New testcase. * g++.dg/gomp/target-3.C: New testcase. * g++.dg/gomp/target-lambda-1.C: New testcase. * g++.dg/gomp/target-lambda-2.C: New testcase. * g++.dg/gomp/target-this-1.C: New testcase. * g++.dg/gomp/target-this-2.C: New testcase. * g++.dg/gomp/target-this-3.C: New testcase. * g++.dg/gomp/target-this-4.C: New testcase. * g++.dg/gomp/target-this-5.C: New testcase. * g++.dg/gomp/this-2.C: Adjust testcase. include/ChangeLog: * gomp-constants.h (enum gomp_map_kind): Add GOMP_MAP_ATTACH_ZERO_LENGTH_ARRAY_SECTION, and GOMP_MAP_POINTER_TO_ZERO_LENGTH_ARRAY_SECTION map kinds. (GOMP_MAP_POINTER_P): Include GOMP_MAP_POINTER_TO_ZERO_LENGTH_ARRAY_SECTION. libgomp/ChangeLog: * libgomp.h (gomp_attach_pointer): Add bool parameter. * oacc-mem.c (acc_attach_async): Update call to gomp_attach_pointer. (goacc_enter_data_internal): Likewise. * target.c (gomp_map_vars_existing): Update assert condition to include GOMP_MAP_ATTACH_ZERO_LENGTH_ARRAY_SECTION. (gomp_map_pointer): Add 'bool allow_zero_length_array_sections' parameter, add support for mapping a pointer with NULL target. (gomp_attach_pointer): Add 'bool allow_zero_length_array_sections' parameter, add support for attaching a pointer with NULL target. (gomp_map_vars_internal): Update calls to gomp_map_pointer and gomp_attach_pointer, add handling for GOMP_MAP_ATTACH_ZERO_LENGTH_ARRAY_SECTION, and GOMP_MAP_POINTER_TO_ZERO_LENGTH_ARRAY_SECTION cases. * testsuite/libgomp.c++/target-23.C: New testcase. * testsuite/libgomp.c++/target-lambda-1.C: New testcase. * testsuite/libgomp.c++/target-lambda-2.C: New testcase. * testsuite/libgomp.c++/target-this-1.C: New testcase. * testsuite/libgomp.c++/target-this-2.C: New testcase. * testsuite/libgomp.c++/target-this-3.C: New testcase. * testsuite/libgomp.c++/target-this-4.C: New testcase. * testsuite/libgomp.c++/target-this-5.C: New testcase.
Diffstat (limited to 'gcc/cp/parser.c')
-rw-r--r--gcc/cp/parser.c69
1 files changed, 13 insertions, 56 deletions
diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c
index 55e6a1a..ed36709 100644
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -39315,7 +39315,7 @@ cp_parser_omp_clause_map (cp_parser *parser, tree list)
}
nlist = cp_parser_omp_var_list_no_open (parser, OMP_CLAUSE_MAP, list,
- NULL);
+ NULL, true);
for (c = nlist; c != list; c = OMP_CLAUSE_CHAIN (c))
OMP_CLAUSE_SET_MAP_KIND (c, kind);
@@ -44105,8 +44105,6 @@ static bool
cp_parser_omp_target (cp_parser *parser, cp_token *pragma_tok,
enum pragma_context context, bool *if_p)
{
- tree *pc = NULL, stmt;
-
if (flag_openmp)
omp_requires_mask
= (enum omp_requires) (omp_requires_mask | OMP_REQUIRES_TARGET_USED);
@@ -44211,16 +44209,10 @@ cp_parser_omp_target (cp_parser *parser, cp_token *pragma_tok,
= cclauses[C_OMP_CLAUSE_SPLIT_TARGET];
cclauses[C_OMP_CLAUSE_SPLIT_TARGET] = tc;
}
- tree stmt = make_node (OMP_TARGET);
- TREE_TYPE (stmt) = void_type_node;
- OMP_TARGET_CLAUSES (stmt) = cclauses[C_OMP_CLAUSE_SPLIT_TARGET];
- c_omp_adjust_map_clauses (OMP_TARGET_CLAUSES (stmt), true);
- OMP_TARGET_BODY (stmt) = body;
- OMP_TARGET_COMBINED (stmt) = 1;
- SET_EXPR_LOCATION (stmt, pragma_tok->location);
- add_stmt (stmt);
- pc = &OMP_TARGET_CLAUSES (stmt);
- goto check_clauses;
+ c_omp_adjust_map_clauses (cclauses[C_OMP_CLAUSE_SPLIT_TARGET], true);
+ finish_omp_target (pragma_tok->location,
+ cclauses[C_OMP_CLAUSE_SPLIT_TARGET], body, true);
+ return true;
}
else if (!flag_openmp) /* flag_openmp_simd */
{
@@ -44255,13 +44247,10 @@ cp_parser_omp_target (cp_parser *parser, cp_token *pragma_tok,
return false;
}
- stmt = make_node (OMP_TARGET);
- TREE_TYPE (stmt) = void_type_node;
-
- OMP_TARGET_CLAUSES (stmt)
- = cp_parser_omp_all_clauses (parser, OMP_TARGET_CLAUSE_MASK,
- "#pragma omp target", pragma_tok, false);
- for (tree c = OMP_TARGET_CLAUSES (stmt); c; c = OMP_CLAUSE_CHAIN (c))
+ tree clauses = cp_parser_omp_all_clauses (parser, OMP_TARGET_CLAUSE_MASK,
+ "#pragma omp target", pragma_tok,
+ false);
+ for (tree c = clauses; c; c = OMP_CLAUSE_CHAIN (c))
if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_IN_REDUCTION)
{
tree nc = build_omp_clause (OMP_CLAUSE_LOCATION (c), OMP_CLAUSE_MAP);
@@ -44270,45 +44259,13 @@ cp_parser_omp_target (cp_parser *parser, cp_token *pragma_tok,
OMP_CLAUSE_CHAIN (nc) = OMP_CLAUSE_CHAIN (c);
OMP_CLAUSE_CHAIN (c) = nc;
}
- OMP_TARGET_CLAUSES (stmt)
- = finish_omp_clauses (OMP_TARGET_CLAUSES (stmt), C_ORT_OMP_TARGET);
- c_omp_adjust_map_clauses (OMP_TARGET_CLAUSES (stmt), true);
+ clauses = finish_omp_clauses (clauses, C_ORT_OMP_TARGET);
- pc = &OMP_TARGET_CLAUSES (stmt);
+ c_omp_adjust_map_clauses (clauses, true);
keep_next_level (true);
- OMP_TARGET_BODY (stmt) = cp_parser_omp_structured_block (parser, if_p);
-
- SET_EXPR_LOCATION (stmt, pragma_tok->location);
- add_stmt (stmt);
+ tree body = cp_parser_omp_structured_block (parser, if_p);
-check_clauses:
- while (*pc)
- {
- if (OMP_CLAUSE_CODE (*pc) == OMP_CLAUSE_MAP)
- switch (OMP_CLAUSE_MAP_KIND (*pc))
- {
- case GOMP_MAP_TO:
- case GOMP_MAP_ALWAYS_TO:
- case GOMP_MAP_FROM:
- case GOMP_MAP_ALWAYS_FROM:
- case GOMP_MAP_TOFROM:
- case GOMP_MAP_ALWAYS_TOFROM:
- case GOMP_MAP_ALLOC:
- case GOMP_MAP_FIRSTPRIVATE_POINTER:
- case GOMP_MAP_FIRSTPRIVATE_REFERENCE:
- case GOMP_MAP_ALWAYS_POINTER:
- case GOMP_MAP_ATTACH_DETACH:
- break;
- default:
- error_at (OMP_CLAUSE_LOCATION (*pc),
- "%<#pragma omp target%> with map-type other "
- "than %<to%>, %<from%>, %<tofrom%> or %<alloc%> "
- "on %<map%> clause");
- *pc = OMP_CLAUSE_CHAIN (*pc);
- continue;
- }
- pc = &OMP_CLAUSE_CHAIN (*pc);
- }
+ finish_omp_target (pragma_tok->location, clauses, body, false);
return true;
}
generated by cgit v1.1 at 2025-04-27 01:04:49 +0000