From mboxrd@z Thu Jan 1 00:00:00 1970 Return-Path: Received: by sourceware.org (Postfix, from userid 1729) id 4C0403896C1C; Thu, 13 May 2021 16:20:07 +0000 (GMT) DKIM-Filter: OpenDKIM Filter v2.11.0 sourceware.org 4C0403896C1C Content-Type: text/plain; charset="us-ascii" MIME-Version: 1.0 Content-Transfer-Encoding: 7bit From: Kwok Yeung To: gcc-cvs@gcc.gnu.org Subject: [gcc/devel/omp/gcc-11] Fix template case of non-static member access inside member functions X-Act-Checkin: gcc X-Git-Author: Chung-Lin Tang X-Git-Refname: refs/heads/devel/omp/gcc-11 X-Git-Oldrev: 60040dfff937acf02e44eb17902ea2d162f4d2ea X-Git-Newrev: 56314e382995ebbf35f77b9a2542b5411b9ae755 Message-Id: <20210513162007.4C0403896C1C@sourceware.org> Date: Thu, 13 May 2021 16:20:07 +0000 (GMT) X-BeenThere: gcc-cvs@gcc.gnu.org X-Mailman-Version: 2.1.29 Precedence: list List-Id: Gcc-cvs mailing list List-Unsubscribe: , List-Archive: List-Help: List-Subscribe: , X-List-Received-Date: Thu, 13 May 2021 16:20:07 -0000 https://gcc.gnu.org/g:56314e382995ebbf35f77b9a2542b5411b9ae755 commit 56314e382995ebbf35f77b9a2542b5411b9ae755 Author: Chung-Lin Tang Date: Thu Mar 11 00:31:08 2021 -0800 Fix template case of non-static member access inside member functions Prior patches for C++ non-static member access had problems under template classes, due to re-calling of finish_omp_clauses after finish_omp_target created the implicit maps required, but not of allowed form in finish_omp_clauses. This patch solves this by slightly relaxing the allowed expressions in finish_omp_clauses. 2021-03-11 Chung-Lin Tang gcc/cp/ChangeLog: * semantics.c (finish_omp_clauses): Adjustments to allow '*ptr' and 'ptr->member' cases in map clausess. (finish_omp_target): Use INDIRECT_REF instead of MEM_REF in created clauses, add processing_template_decl handling. gcc/ChangeLog: * gimplify.c (gimplify_scan_omp_clauses): 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. gcc/testsuite/ChangeLog: * g++.dg/gomp/target-this-3.C: Adjust scan test. * g++.dg/gomp/target-this-4.C: Likewise. * g++.dg/gomp/target-this-5.C: New test. libgomp/ChangeLog: * testsuite/libgomp.c++/target-this-5.C: New test. Diff: --- gcc/cp/semantics.c | 45 +++++++++++++++++++++------ gcc/gimplify.c | 19 +++++++++++ gcc/testsuite/g++.dg/gomp/target-this-3.C | 2 +- gcc/testsuite/g++.dg/gomp/target-this-4.C | 2 +- gcc/testsuite/g++.dg/gomp/target-this-5.C | 34 ++++++++++++++++++++ libgomp/testsuite/libgomp.c++/target-this-5.C | 30 ++++++++++++++++++ 6 files changed, 120 insertions(+), 12 deletions(-) diff --git a/gcc/cp/semantics.c b/gcc/cp/semantics.c index d37e231641c..46eb5807eb9 100644 --- a/gcc/cp/semantics.c +++ b/gcc/cp/semantics.c @@ -6600,6 +6600,7 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort) bool order_seen = false; bool schedule_seen = false; bool oacc_async = false; + bool indirect_ref_p = false; bool indir_component_ref_p = false; tree last_iterators = NULL_TREE; bool last_iterators_remove = false; @@ -7784,6 +7785,14 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort) indir_component_ref_p = true; STRIP_NOPS (t); } + indirect_ref_p = false; + if ((ort == C_ORT_ACC || ort == C_ORT_OMP) + && INDIRECT_REF_P (t)) + { + t = TREE_OPERAND (t, 0); + indirect_ref_p = true; + STRIP_NOPS (t); + } if (TREE_CODE (t) == COMPONENT_REF && ((ort & C_ORT_OMP_DECLARE_SIMD) == C_ORT_OMP || ort == C_ORT_ACC) @@ -7819,6 +7828,12 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort) break; } t = TREE_OPERAND (t, 0); + if (INDIRECT_REF_P (t)) + { + t = TREE_OPERAND (t, 0); + indir_component_ref_p = true; + STRIP_NOPS (t); + } } if (remove) break; @@ -7882,6 +7897,7 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort) || (OMP_CLAUSE_MAP_KIND (c) != GOMP_MAP_FIRSTPRIVATE_POINTER)) && !indir_component_ref_p + && !indirect_ref_p && !cxx_mark_addressable (t)) remove = true; else if (!(OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP @@ -7966,7 +7982,8 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort) } else { - bitmap_set_bit (&map_head, DECL_UID (t)); + if (!indirect_ref_p && !indir_component_ref_p) + bitmap_set_bit (&map_head, DECL_UID (t)); if (t != OMP_CLAUSE_DECL (c) && TREE_CODE (OMP_CLAUSE_DECL (c)) == COMPONENT_REF) bitmap_set_bit (&map_field_head, DECL_UID (t)); @@ -9027,9 +9044,12 @@ finish_omp_target (location_t loc, tree clauses, tree body, bool combined_p) tree closure = DECL_ARGUMENTS (current_function_decl); tree c = build_omp_clause (loc, OMP_CLAUSE_MAP); OMP_CLAUSE_SET_MAP_KIND (c, GOMP_MAP_TO); - OMP_CLAUSE_DECL (c) = build_simple_mem_ref (closure); + OMP_CLAUSE_DECL (c) + = build_indirect_ref (loc, closure, RO_UNARY_STAR); OMP_CLAUSE_SIZE (c) - = TYPE_SIZE_UNIT (TREE_TYPE (TREE_TYPE (closure))); + = (processing_template_decl + ? NULL_TREE + : TYPE_SIZE_UNIT (TREE_TYPE (TREE_TYPE (closure)))); tree c2 = build_omp_clause (loc, OMP_CLAUSE_MAP); OMP_CLAUSE_SET_MAP_KIND (c2, GOMP_MAP_FIRSTPRIVATE_POINTER); @@ -9049,7 +9069,8 @@ finish_omp_target (location_t loc, tree clauses, tree body, bool combined_p) /* Transform *this into *__closure->this in maps. */ tree this_map = *explicit_this_deref_map; OMP_CLAUSE_DECL (this_map) - = build_simple_mem_ref (omp_target_this_expr); + = build_indirect_ref (loc, omp_target_this_expr, RO_UNARY_STAR); + tree nc = OMP_CLAUSE_CHAIN (this_map); gcc_assert (OMP_CLAUSE_CODE (nc) == OMP_CLAUSE_MAP && (OMP_CLAUSE_MAP_KIND (nc) @@ -9069,9 +9090,11 @@ finish_omp_target (location_t loc, tree clauses, tree body, bool combined_p) tree c3 = build_omp_clause (loc, OMP_CLAUSE_MAP); OMP_CLAUSE_SET_MAP_KIND (c3, GOMP_MAP_TOFROM); OMP_CLAUSE_DECL (c3) - = build_simple_mem_ref (omp_target_this_expr); + = build_indirect_ref (loc, omp_target_this_expr, RO_UNARY_STAR); OMP_CLAUSE_SIZE (c3) - = TYPE_SIZE_UNIT (TREE_TYPE (TREE_TYPE (omp_target_this_expr))); + = (processing_template_decl + ? NULL_TREE + : TYPE_SIZE_UNIT (TREE_TYPE (TREE_TYPE (omp_target_this_expr)))); tree c4 = build_omp_clause (loc, OMP_CLAUSE_MAP); OMP_CLAUSE_SET_MAP_KIND (c4, GOMP_MAP_ALWAYS_POINTER); @@ -9093,9 +9116,12 @@ finish_omp_target (location_t loc, tree clauses, tree body, bool combined_p) { tree c = build_omp_clause (loc, OMP_CLAUSE_MAP); OMP_CLAUSE_SET_MAP_KIND (c, GOMP_MAP_TOFROM); - OMP_CLAUSE_DECL (c) = build_simple_mem_ref (omp_target_this_expr); + OMP_CLAUSE_DECL (c) + = build_indirect_ref (loc, omp_target_this_expr, RO_UNARY_STAR); OMP_CLAUSE_SIZE (c) - = TYPE_SIZE_UNIT (TREE_TYPE (TREE_TYPE (omp_target_this_expr))); + = (processing_template_decl + ? NULL_TREE + : TYPE_SIZE_UNIT (TREE_TYPE (TREE_TYPE (omp_target_this_expr)))); tree c2 = build_omp_clause (loc, OMP_CLAUSE_MAP); OMP_CLAUSE_SET_MAP_KIND (c2, GOMP_MAP_FIRSTPRIVATE_POINTER); @@ -9177,8 +9203,7 @@ finish_omp_target (location_t loc, tree clauses, tree body, bool combined_p) tree c = build_omp_clause (loc, OMP_CLAUSE_MAP); OMP_CLAUSE_SET_MAP_KIND (c, GOMP_MAP_ALLOC); OMP_CLAUSE_DECL (c) - = build2 (MEM_REF, char_type_node, ptr_member, - build_int_cst (build_pointer_type (char_type_node), 0)); + = build_indirect_ref (loc, ptr_member, RO_UNARY_STAR); OMP_CLAUSE_SIZE (c) = size_zero_node; OMP_CLAUSE_MAP_MAYBE_ZERO_LENGTH_ARRAY_SECTION (c) = 1; diff --git a/gcc/gimplify.c b/gcc/gimplify.c index 03fbe461bd5..0dd8e7d71b4 100644 --- a/gcc/gimplify.c +++ b/gcc/gimplify.c @@ -9178,6 +9178,7 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p, { indir_p = true; decl = TREE_OPERAND (decl, 0); + STRIP_NOPS (decl); } if (TREE_CODE (decl) == INDIRECT_REF && DECL_P (TREE_OPERAND (decl, 0)) @@ -9589,6 +9590,24 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p, break; } + /* If this was of the form map(*pointer_to_struct), then the + 'pointer_to_struct' DECL should be considered deref'ed. */ + if ((OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_ALLOC + || GOMP_MAP_COPY_TO_P (OMP_CLAUSE_MAP_KIND (c)) + || GOMP_MAP_COPY_FROM_P (OMP_CLAUSE_MAP_KIND (c))) + && INDIRECT_REF_P (orig_decl) + && DECL_P (TREE_OPERAND (orig_decl, 0)) + && TREE_CODE (TREE_TYPE (orig_decl)) == RECORD_TYPE) + { + tree ptr = TREE_OPERAND (orig_decl, 0); + if (!struct_deref_set || !struct_deref_set->contains (ptr)) + { + if (!struct_deref_set) + struct_deref_set = new hash_set (); + struct_deref_set->add (ptr); + } + } + if (!remove && OMP_CLAUSE_MAP_KIND (c) != GOMP_MAP_ALWAYS_POINTER && OMP_CLAUSE_MAP_KIND (c) != GOMP_MAP_ATTACH_DETACH diff --git a/gcc/testsuite/g++.dg/gomp/target-this-3.C b/gcc/testsuite/g++.dg/gomp/target-this-3.C index a450b3723e5..08568f9284c 100644 --- a/gcc/testsuite/g++.dg/gomp/target-this-3.C +++ b/gcc/testsuite/g++.dg/gomp/target-this-3.C @@ -102,4 +102,4 @@ int main (void) /* { dg-final { scan-tree-dump {#pragma omp target num_teams.* map\(tofrom:\*this \[len: [0-9]+\]\) map\(firstprivate:this \[pointer assign, bias: 0\]\) map\(alloc:\*_[0-9]+ \[pointer assign, zero-length array section, bias: 0\]\) map\(attach:this->refptr \[bias: 0\]\) map\(from:mapped \[len: [0-9]+\]\) map\(alloc:\*_[0-9+] \[len: 0\]\) firstprivate\(n\)} "gimple" } } */ -/* { dg-final { scan-tree-dump {#pragma omp target num_teams.* map\(tofrom:\*this \[len: [0-9]+\]\) map\(firstprivate:this \[pointer assign, bias: 0\]\) map\(attach_zero_length_array_section:this->ptr \[bias: 0\]\) map\(from:mapped \[len: [0-9]+\]\) map\(alloc:MEM.* \[len: 0\]\) firstprivate\(n\)} "gimple" } } */ +/* { dg-final { scan-tree-dump {#pragma omp target num_teams.* map\(tofrom:\*this \[len: [0-9]+\]\) map\(firstprivate:this \[pointer assign, bias: 0\]\) map\(attach_zero_length_array_section:this->ptr \[bias: 0\]\) map\(from:mapped \[len: [0-9]+\]\) map\(alloc:\*_[0-9]+ \[len: 0\]\) firstprivate\(n\)} "gimple" } } */ diff --git a/gcc/testsuite/g++.dg/gomp/target-this-4.C b/gcc/testsuite/g++.dg/gomp/target-this-4.C index af23cbb9023..3b2d5811350 100644 --- a/gcc/testsuite/g++.dg/gomp/target-this-4.C +++ b/gcc/testsuite/g++.dg/gomp/target-this-4.C @@ -102,6 +102,6 @@ int main (void) return 0; } -/* { dg-final { scan-tree-dump {#pragma omp target num_teams.* map\(to:\*__closure \[len: [0-9]+\]\) map\(firstprivate:__closure \[pointer assign, bias: 0\]\) map\(struct:\*__closure \[len: 1\]\) map\(alloc:__closure->__this \[len: [0-9]+\]\) map\(tofrom:\*_[0-9]+ \[len: [0-9]+\]\) map\(always_pointer:__closure->__this \[pointer assign, bias: 0\]\) map\(attach_zero_length_array_section:_[0-9]+->ptr \[bias: 0\]\) map\(from:mapped \[len: 1\]\) map\(alloc:MEM.* \[len: 0\]\) firstprivate\(n\) map\(alloc:MEM.* \[len: 0\]\) map\(firstprivate:this \[pointer assign, bias: 0\]\)} "gimple" } } */ +/* { dg-final { scan-tree-dump {#pragma omp target num_teams.* map\(to:\*__closure \[len: [0-9]+\]\) map\(firstprivate:__closure \[pointer assign, bias: 0\]\) map\(struct:\*__closure \[len: 1\]\) map\(alloc:__closure->__this \[len: [0-9]+\]\) map\(tofrom:\*_[0-9]+ \[len: [0-9]+\]\) map\(always_pointer:__closure->__this \[pointer assign, bias: 0\]\) map\(attach_zero_length_array_section:_[0-9]+->ptr \[bias: 0\]\) map\(from:mapped \[len: 1\]\) map\(alloc:\*_[0-9]+ \[len: 0\]\) firstprivate\(n\) map\(alloc:MEM.* \[len: 0\]\) map\(firstprivate:this \[pointer assign, bias: 0\]\)} "gimple" } } */ /* { dg-final { scan-tree-dump {#pragma omp target num_teams.* map\(to:\*__closure \[len: [0-9]+\]\) map\(firstprivate:__closure \[pointer assign, bias: 0\]\) map\(struct:\*__closure \[len: 1\]\) map\(alloc:__closure->__this \[len: [0-9]+\]\) map\(tofrom:\*_[0-9]+ \[len: [0-9]+\]\) map\(always_pointer:__closure->__this \[pointer assign, bias: 0\]\) map\(alloc:\*_[0-9]+ \[pointer assign, zero-length array section, bias: 0\]\) map\(attach:_[0-9]+->refptr \[bias: 0\]\) map\(from:mapped \[len: [0-9]+\]\) map\(alloc:\*_[0-9]+ \[len: 0\]\) firstprivate\(n\) map\(alloc:MEM.* \[len: 0\]\) map\(firstprivate:this \[pointer assign, bias: 0\]\)} "gimple" } } */ diff --git a/gcc/testsuite/g++.dg/gomp/target-this-5.C b/gcc/testsuite/g++.dg/gomp/target-this-5.C new file mode 100644 index 00000000000..a9ac74bcf1f --- /dev/null +++ b/gcc/testsuite/g++.dg/gomp/target-this-5.C @@ -0,0 +1,34 @@ +// { dg-do compile } +// { dg-additional-options "-fdump-tree-gimple" } +extern "C" void abort (); + +template +struct S +{ + T a, b, c, d; + + T sum (void) + { + T val = 0; + val += a + b + this->c + this->d; + return val; + } + + T sum_offload (void) + { + T val = 0; + #pragma omp target map(val) + val += a + b + this->c + this->d; + return val; + } +}; + +int main (void) +{ + S s = { 1, 2, 3, 4 }; + if (s.sum () != s.sum_offload ()) + abort (); + return 0; +} + +/* { dg-final { scan-tree-dump {map\(tofrom:\*this \[len: [0-9]+\]\) map\(firstprivate:this \[pointer assign, bias: 0\]\)} "gimple" } } */ diff --git a/libgomp/testsuite/libgomp.c++/target-this-5.C b/libgomp/testsuite/libgomp.c++/target-this-5.C new file mode 100644 index 00000000000..e71c566687d --- /dev/null +++ b/libgomp/testsuite/libgomp.c++/target-this-5.C @@ -0,0 +1,30 @@ +extern "C" void abort (); + +template +struct S +{ + T a, b, c, d; + + T sum (void) + { + T val = 0; + val += a + b + this->c + this->d; + return val; + } + + T sum_offload (void) + { + T val = 0; + #pragma omp target map(val) + val += a + b + this->c + this->d; + return val; + } +}; + +int main (void) +{ + S s = { 1, 2, 3, 4 }; + if (s.sum () != s.sum_offload ()) + abort (); + return 0; +}