OpenMP: Ignore side-effects when finding struct comps [PR108545] With volatile, two 'x.data' comp refs aren't regarded as identical, causing that the two items in the first map of map(to:x.a, x.a.data) map(pset: x.a.data) end up in separate 'map(struct:x)', which will cause a later ICE. Solution: Ignore side effects when checking the operands in the hash for being equal. (Do so by creating a variant of tree_operand_hash that calls operand_equal_p with OEP_MATCH_SIDE_EFFECTS.) gcc/ChangeLog: PR middle-end/108545 * gimplify.cc (struct tree_operand_sideeff_hash): New. omp_index_mapping_groups_1, omp_index_mapping_groups, omp_reindex_mapping_groups, omp_mapped_by_containing_struct, omp_tsort_mapping_groups_1, omp_tsort_mapping_groups, oacc_resolve_clause_dependencies, omp_build_struct_sibling_lists, gimplify_scan_omp_clauses): Use tree_operand_sideeff_hash instead of tree_operand_hash. gcc/testsuite/ChangeLog: PR middle-end/108545 * c-c++-common/gomp/map-8.c: New test. * gfortran.dg/gomp/map-9.f90: New test. gcc/gimplify.cc | 55 ++++++++++++++++++++++---------- gcc/testsuite/c-c++-common/gomp/map-8.c | 19 +++++++++++ gcc/testsuite/gfortran.dg/gomp/map-9.f90 | 13 ++++++++ 3 files changed, 70 insertions(+), 17 deletions(-) diff --git a/gcc/gimplify.cc b/gcc/gimplify.cc index 96845154a92..3c0bb1c987a 100644 --- a/gcc/gimplify.cc +++ b/gcc/gimplify.cc @@ -8958,6 +8958,28 @@ enum omp_tsort_mark { PERMANENT }; +/* Hash for trees based on operand_equal_p. + Like tree_operand_hash but accepts side effects. */ +struct tree_operand_sideeff_hash : ggc_ptr_hash +{ + static inline hashval_t hash (const value_type &); + static inline bool equal (const value_type &, + const compare_type &); +}; + +inline hashval_t +tree_operand_sideeff_hash::hash (const value_type &t) +{ + return iterative_hash_expr (t, 0); +} + +inline bool +tree_operand_sideeff_hash::equal (const value_type &t1, + const compare_type &t2) +{ + return operand_equal_p (t1, t2, OEP_MATCH_SIDE_EFFECTS); +} + /* A group of OMP_CLAUSE_MAP nodes that correspond to a single "map" clause. */ @@ -9400,10 +9422,10 @@ omp_group_base (omp_mapping_group *grp, unsigned int *chained, } /* Given a vector of omp_mapping_groups, build a hash table so we can look up - nodes by tree_operand_hash. */ + nodes by tree_operand_sideeff_hash. */ static void -omp_index_mapping_groups_1 (hash_map *grpmap, vec *groups, tree reindex_sentinel) @@ -9432,7 +9454,6 @@ omp_index_mapping_groups_1 (hash_map * +static hash_map * omp_index_mapping_groups (vec *groups) { - hash_map *grpmap - = new hash_map; + hash_map *grpmap + = new hash_map; omp_index_mapping_groups_1 (grpmap, groups, NULL_TREE); @@ -9502,14 +9523,14 @@ omp_index_mapping_groups (vec *groups) so, we can do the reindex operation in two parts, on the processed and then the unprocessed halves of the list. */ -static hash_map * +static hash_map * omp_reindex_mapping_groups (tree *list_p, vec *groups, vec *processed_groups, tree sentinel) { - hash_map *grpmap - = new hash_map; + hash_map *grpmap + = new hash_map; processed_groups->truncate (0); @@ -9550,7 +9571,7 @@ omp_containing_struct (tree expr) that maps that structure, if present. */ static bool -omp_mapped_by_containing_struct (hash_map *grpmap, tree decl, omp_mapping_group **mapped_by_group) @@ -9590,7 +9611,7 @@ omp_mapped_by_containing_struct (hash_map *groups, - hash_map + hash_map *grpmap, omp_mapping_group *grp) { @@ -9670,7 +9691,7 @@ omp_tsort_mapping_groups_1 (omp_mapping_group ***outlist, static omp_mapping_group * omp_tsort_mapping_groups (vec *groups, - hash_map + hash_map *grpmap) { omp_mapping_group *grp, *outlist = NULL, **cursor; @@ -9986,7 +10007,7 @@ omp_check_mapping_compatibility (location_t loc, void oacc_resolve_clause_dependencies (vec *groups, - hash_map *grpmap) { int i; @@ -10520,8 +10541,8 @@ static bool omp_build_struct_sibling_lists (enum tree_code code, enum omp_region_type region_type, vec *groups, - hash_map - **grpmap, + hash_map **grpmap, tree *list_p) { unsigned i; @@ -10747,7 +10768,7 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p, groups = omp_gather_mapping_groups (list_p); if (groups) { - hash_map *grpmap; + hash_map *grpmap; grpmap = omp_index_mapping_groups (groups); omp_build_struct_sibling_lists (code, region_type, groups, &grpmap, @@ -10783,7 +10804,7 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p, groups = omp_gather_mapping_groups (list_p); if (groups) { - hash_map *grpmap; + hash_map *grpmap; grpmap = omp_index_mapping_groups (groups); oacc_resolve_clause_dependencies (groups, grpmap); diff --git a/gcc/testsuite/c-c++-common/gomp/map-8.c b/gcc/testsuite/c-c++-common/gomp/map-8.c new file mode 100644 index 00000000000..cc7dac8bd0f --- /dev/null +++ b/gcc/testsuite/c-c++-common/gomp/map-8.c @@ -0,0 +1,19 @@ +/* { dg-additional-options "-fdump-tree-omplower" } */ + +/* PR fortran/108545 */ + +/* { dg-final { scan-tree-dump "#pragma omp target enter data map\\(struct:my_struct \\\[len: 1\\\]\\) map\\(to:my_struct.u \\\[len: \[0-9\]+\\\]\\)" "omplower" } } */ +/* { dg-final { scan-tree-dump "#pragma omp target enter data map\\(to:my_struct3 \\\[len: \[0-9\]+\\\]\\)" "omplower" } } */ + + +volatile struct t { + struct t2 { int *a; int c; } u; + int b; +} my_struct; +volatile struct t3 { int *a; int c; } my_struct3; + +void f() +{ + #pragma omp target enter data map(to:my_struct.u) map(to:my_struct.u.a) + #pragma omp target enter data map(to:my_struct3) map(to:my_struct3.a) +} diff --git a/gcc/testsuite/gfortran.dg/gomp/map-9.f90 b/gcc/testsuite/gfortran.dg/gomp/map-9.f90 new file mode 100644 index 00000000000..9e7b811c8af --- /dev/null +++ b/gcc/testsuite/gfortran.dg/gomp/map-9.f90 @@ -0,0 +1,13 @@ +! { dg-additional-options "-fdump-tree-omplower" } + +! PR fortran/108545 + +! { dg-final { scan-tree-dump "#pragma omp target enter data map\\(struct:x \\\[len: 1\\\]\\) map\\(to:x.a \\\[len: \[0-9\]+\\\]\\) map\\(to:MEM \\\[\\(integer\\(kind=4\\)\\\[0:\\\] \\*\\)_\[0-9\]+] \\\[len: _\[0-9\]+\\\]\\) map\\(always_pointer:x.a.data \\\[pointer assign, bias: 0\\\]\\)" "omplower" } } + +program p + type t + integer, pointer :: a(:) + end type + type(t), volatile :: x + !$omp target enter data map(to: x%a) +end