[gcc/devel/omp/gcc-10] Arrow operator handling for C front-end in OpenMP map clauses
Chung-Lin Tang
cltang@gcc.gnu.org
Mon Mar 8 07:57:34 GMT 2021
https://gcc.gnu.org/g:08caada8efd8f35db634647bbda6091fb667b00d
commit 08caada8efd8f35db634647bbda6091fb667b00d
Author: Chung-Lin Tang <cltang@codesourcery.com>
Date: Mon Mar 8 15:56:52 2021 +0800
Arrow operator handling for C front-end in OpenMP map clauses
This patch merges some of the equivalent changes already done for the C++
front-end to the C parts.
2021-03-08 Chung-Lin Tang <cltang@codesourcery.com>
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 (gimplify_scan_omp_clauses): 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.
gcc/testsuite/ChangeLog:
* gcc.dg/gomp/target-3.c: New test.
Diff:
---
gcc/c/c-parser.c | 3 ++-
gcc/c/c-typeck.c | 22 +++++++++++++++++++
gcc/gimplify.c | 41 ++++++++++++++++++++++++++++++++++--
gcc/testsuite/gcc.dg/gomp/target-3.c | 16 ++++++++++++++
4 files changed, 79 insertions(+), 3 deletions(-)
diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c
index fae597128e9..0a6aee439f6 100644
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -15700,7 +15700,8 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
}
}
- nl = c_parser_omp_variable_list (parser, clause_loc, OMP_CLAUSE_MAP, list);
+ nl = c_parser_omp_variable_list (parser, clause_loc, OMP_CLAUSE_MAP, list,
+ C_ORT_OMP, true);
for (c = nl; c != list; c = OMP_CLAUSE_CHAIN (c))
OMP_CLAUSE_SET_MAP_KIND (c, kind);
diff --git a/gcc/c/c-typeck.c b/gcc/c/c-typeck.c
index 6af19766324..7c887a80ce9 100644
--- a/gcc/c/c-typeck.c
+++ b/gcc/c/c-typeck.c
@@ -12917,6 +12917,12 @@ handle_omp_array_sections_1 (tree c, tree t, vec<tree> &types,
return error_mark_node;
}
t = TREE_OPERAND (t, 0);
+ if ((ort == C_ORT_ACC || ort == C_ORT_OMP)
+ && TREE_CODE (t) == MEM_REF)
+ {
+ t = TREE_OPERAND (t, 0);
+ STRIP_NOPS (t);
+ }
if (ort == C_ORT_ACC && TREE_CODE (t) == MEM_REF)
{
if (maybe_ne (mem_ref_offset (t), 0))
@@ -13778,6 +13784,7 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
tree ordered_clause = NULL_TREE;
tree schedule_clause = NULL_TREE;
bool oacc_async = false;
+ bool indir_component_ref_p = false;
tree last_iterators = NULL_TREE;
bool last_iterators_remove = false;
tree *nogroup_seen = NULL;
@@ -14505,6 +14512,11 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
{
while (TREE_CODE (t) == COMPONENT_REF)
t = TREE_OPERAND (t, 0);
+ if (TREE_CODE (t) == MEM_REF)
+ {
+ t = TREE_OPERAND (t, 0);
+ STRIP_NOPS (t);
+ }
if (bitmap_bit_p (&map_field_head, DECL_UID (t)))
break;
if (bitmap_bit_p (&map_head, DECL_UID (t)))
@@ -14561,6 +14573,15 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
bias) to zero here, so it is not set erroneously to the pointer
size later on in gimplify.c. */
OMP_CLAUSE_SIZE (c) = size_zero_node;
+ indir_component_ref_p = false;
+ if ((ort == C_ORT_ACC || ort == C_ORT_OMP)
+ && TREE_CODE (t) == COMPONENT_REF
+ && TREE_CODE (TREE_OPERAND (t, 0)) == MEM_REF)
+ {
+ t = TREE_OPERAND (TREE_OPERAND (t, 0), 0);
+ indir_component_ref_p = true;
+ STRIP_NOPS (t);
+ }
if (TREE_CODE (t) == COMPONENT_REF
&& OMP_CLAUSE_CODE (c) != OMP_CLAUSE__CACHE_)
{
@@ -14633,6 +14654,7 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
else if ((OMP_CLAUSE_CODE (c) != OMP_CLAUSE_MAP
|| (OMP_CLAUSE_MAP_KIND (c)
!= GOMP_MAP_FIRSTPRIVATE_POINTER))
+ && !indir_component_ref_p
&& !c_mark_addressable (t))
remove = true;
else if (!(OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP
diff --git a/gcc/gimplify.c b/gcc/gimplify.c
index 0a6ceb584b1..ff44d26c98e 100644
--- a/gcc/gimplify.c
+++ b/gcc/gimplify.c
@@ -9197,7 +9197,9 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
&& OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_ATTACH_DETACH)
OMP_CLAUSE_SET_MAP_KIND (c, GOMP_MAP_ALWAYS_POINTER);
if ((DECL_P (decl)
- || (component_ref_p && INDIRECT_REF_P (decl)))
+ || (component_ref_p
+ && (INDIRECT_REF_P (decl)
+ || TREE_CODE (decl) == MEM_REF)))
&& OMP_CLAUSE_MAP_KIND (c) != GOMP_MAP_TO_PSET
&& OMP_CLAUSE_MAP_KIND (c) != GOMP_MAP_ATTACH
&& OMP_CLAUSE_MAP_KIND (c) != GOMP_MAP_DETACH
@@ -9298,7 +9300,18 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
if (base_ref)
OMP_CLAUSE_DECL (l) = unshare_expr (base_ref);
else
- OMP_CLAUSE_DECL (l) = decl;
+ {
+ OMP_CLAUSE_DECL (l) = unshare_expr (decl);
+ if (!DECL_P (OMP_CLAUSE_DECL (l))
+ && (gimplify_expr (&OMP_CLAUSE_DECL (l),
+ pre_p, NULL, is_gimple_lvalue,
+ fb_lvalue)
+ == GS_ERROR))
+ {
+ remove = true;
+ break;
+ }
+ }
OMP_CLAUSE_SIZE (l)
= (!attach
? size_int (1)
@@ -9341,6 +9354,30 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
flags |= GOVD_SEEN;
if (has_attachments)
flags |= GOVD_MAP_HAS_ATTACHMENTS;
+
+ /* If this is a *pointer-to-struct expression, make sure a
+ firstprivate map of the base-pointer exists. */
+ if (component_ref_p
+ && ((TREE_CODE (decl) == MEM_REF
+ && integer_zerop (TREE_OPERAND (decl, 1)))
+ || INDIRECT_REF_P (decl))
+ && DECL_P (TREE_OPERAND (decl, 0))
+ && !splay_tree_lookup (ctx->variables,
+ ((splay_tree_key)
+ TREE_OPERAND (decl, 0))))
+ {
+ decl = TREE_OPERAND (decl, 0);
+ tree c2 = build_omp_clause (OMP_CLAUSE_LOCATION (c),
+ OMP_CLAUSE_MAP);
+ enum gomp_map_kind mkind
+ = GOMP_MAP_FIRSTPRIVATE_POINTER;
+ OMP_CLAUSE_SET_MAP_KIND (c2, mkind);
+ OMP_CLAUSE_DECL (c2) = decl;
+ OMP_CLAUSE_SIZE (c2) = size_zero_node;
+ OMP_CLAUSE_CHAIN (c2) = OMP_CLAUSE_CHAIN (c);
+ OMP_CLAUSE_CHAIN (c) = c2;
+ }
+
if (DECL_P (decl))
goto do_add_decl;
}
diff --git a/gcc/testsuite/gcc.dg/gomp/target-3.c b/gcc/testsuite/gcc.dg/gomp/target-3.c
new file mode 100644
index 00000000000..3e7921270c9
--- /dev/null
+++ b/gcc/testsuite/gcc.dg/gomp/target-3.c
@@ -0,0 +1,16 @@
+/* { dg-do compile } */
+/* { dg-options "-fopenmp -fdump-tree-gimple" } */
+
+struct S
+{
+ int a, b;
+};
+
+void foo (struct S *s)
+{
+ #pragma omp target map (alloc: s->a, s->b)
+ ;
+ #pragma omp target enter data map (alloc: s->a, s->b)
+}
+
+/* { dg-final { scan-tree-dump-times "map\\(struct:\\*s \\\[len: 2\\\]\\) map\\(alloc:s->a \\\[len: \[0-9\]+\\\]\\) map\\(alloc:s->b \\\[len: \[0-9\]+\\\]\\)" 2 "gimple" } } */
More information about the Gcc-cvs
mailing list