openmp: Fix up handling of reduction clause on constructs combined with target [PR99928]

The reduction clause should be copied as map (tofrom: ) to combined target if
present, but as we need different handling of array sections between map and reduction,
doing that during gimplification would be harder.

So, this patch adds them during splitting, and similarly to firstprivate adds them
with a new flag that they should be just ignored/removed if an explicit map clause
of the same list item is present.

The exact rules are to be decided in https://github.com/OpenMP/spec/issues/2766
so this patch just implements something that is IMHO reasonable and exact detailed
testcases for the cornercases will follow once it is clarified.

2021-05-28  Jakub Jelinek  <jakub@redhat.com>

	PR middle-end/99928
gcc/
	* tree.h (OMP_CLAUSE_MAP_IMPLICIT): Define.
gcc/c-family/
	* c-omp.c (c_omp_split_clauses): For reduction clause if combined with
	target add a map tofrom clause with OMP_CLAUSE_MAP_IMPLICIT.
gcc/c/
	* c-typeck.c (handle_omp_array_sections): Copy OMP_CLAUSE_MAP_IMPLICIT.
	(c_finish_omp_clauses): Move not just OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT
	marked clauses last, but also OMP_CLAUSE_MAP_IMPLICIT.  Add
	map_firstprivate_head bitmap, set it for GOMP_MAP_FIRSTPRIVATE_POINTER
	maps and silently remove OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT if it is
	present too.  For OMP_CLAUSE_MAP_IMPLICIT silently remove the clause
	if present in map_head, map_field_head or map_firstprivate_head
	bitmaps.
gcc/cp/
	* semantics.c (handle_omp_array_sections): Copy
	OMP_CLAUSE_MAP_IMPLICIT.
	(finish_omp_clauses): Move not just OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT
	marked clauses last, but also OMP_CLAUSE_MAP_IMPLICIT.  Add
	map_firstprivate_head bitmap, set it for GOMP_MAP_FIRSTPRIVATE_POINTER
	maps and silently remove OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT if it is
	present too.  For OMP_CLAUSE_MAP_IMPLICIT silently remove the clause
	if present in map_head, map_field_head or map_firstprivate_head
	bitmaps.
gcc/testsuite/
	* c-c++-common/gomp/pr99928-8.c: Remove all xfails.
	* c-c++-common/gomp/pr99928-9.c: Likewise.
	* c-c++-common/gomp/pr99928-10.c: Likewise.
	* c-c++-common/gomp/pr99928-16.c: New test.

(cherry picked from commit c94424b0ed786ec92b6904da69af8b5243b34fdc)

diff --git a/gcc/ChangeLog.omp b/gcc/ChangeLog.omp
index 5d00320ec6379bdf7811ea75df829ac4d69ff671..041f3fbf19a95f9aa7c570c9c2d463ce2cd0d43e 100644 (file)
--- a/gcc/ChangeLog.omp
+++ b/gcc/ChangeLog.omp
@@ -1,3 +1,10 @@
+2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
+
+       Backported from master:
+       2021-05-28  Jakub Jelinek  <jakub@redhat.com>
+
+       * tree.h (OMP_CLAUSE_MAP_IMPLICIT): Define.
+
 2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
 
        Backported from master:

diff --git a/gcc/c-family/ChangeLog.omp b/gcc/c-family/ChangeLog.omp
index 4e9e98c371d986e78f4802017107c1ae3bb1c94b..f6cdfe189f34c4d90cc660b0d49e71c0e44187a4 100644 (file)
--- a/gcc/c-family/ChangeLog.omp
+++ b/gcc/c-family/ChangeLog.omp
@@ -1,3 +1,11 @@
+2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
+
+       Backported from master:
+       2021-05-28  Jakub Jelinek  <jakub@redhat.com>
+
+       * c-omp.c (c_omp_split_clauses): For reduction clause if combined with
+       target add a map tofrom clause with OMP_CLAUSE_MAP_IMPLICIT.
+
 2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
 
        Backported from master:

diff --git a/gcc/c-family/c-omp.c b/gcc/c-family/c-omp.c
index 23c6674fff5d19341f8ffaaf1b90f6c734ac1ffd..be768391053c5c9f53f3e72e0d80786218dbd338 100644 (file)
--- a/gcc/c-family/c-omp.c
+++ b/gcc/c-family/c-omp.c
@@ -2133,6 +2133,16 @@ c_omp_split_clauses (location_t loc, enum tree_code code,
                        "%<parallel for%>, %<parallel for simd%>");
              OMP_CLAUSE_REDUCTION_INSCAN (clauses) = 0;
            }
+         if ((mask & (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_MAP)) != 0)
+           {
+             c = build_omp_clause (OMP_CLAUSE_LOCATION (clauses),
+                                   OMP_CLAUSE_MAP);
+             OMP_CLAUSE_DECL (c) = OMP_CLAUSE_DECL (clauses);
+             OMP_CLAUSE_SET_MAP_KIND (c, GOMP_MAP_TOFROM);
+             OMP_CLAUSE_MAP_IMPLICIT (c) = 1;
+             OMP_CLAUSE_CHAIN (c) = cclauses[C_OMP_CLAUSE_SPLIT_TARGET];
+             cclauses[C_OMP_CLAUSE_SPLIT_TARGET] = c;
+           }
          if ((mask & (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_SCHEDULE)) != 0)
            {
              if (code == OMP_SIMD)

diff --git a/gcc/c/ChangeLog.omp b/gcc/c/ChangeLog.omp
index e2b66a18472280309e4eb0085754490091ae1e6a..c42f0d8b1145c3e3193707fb6ea1eb37367c8aad 100644 (file)
--- a/gcc/c/ChangeLog.omp
+++ b/gcc/c/ChangeLog.omp
@@ -1,3 +1,17 @@
+2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
+
+       Backported from master:
+       2021-05-28  Jakub Jelinek  <jakub@redhat.com>
+
+       * c-typeck.c (handle_omp_array_sections): Copy OMP_CLAUSE_MAP_IMPLICIT.
+       (c_finish_omp_clauses): Move not just OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT
+       marked clauses last, but also OMP_CLAUSE_MAP_IMPLICIT.  Add
+       map_firstprivate_head bitmap, set it for GOMP_MAP_FIRSTPRIVATE_POINTER
+       maps and silently remove OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT if it is
+       present too.  For OMP_CLAUSE_MAP_IMPLICIT silently remove the clause
+       if present in map_head, map_field_head or map_firstprivate_head
+       bitmaps.
+
 2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
 
        Backported from master:

diff --git a/gcc/c/c-typeck.c b/gcc/c/c-typeck.c
index f8096430d35aeb34a5e45278267c95e887dfbbe3..f5aa8af21bdd096a6808c31bffc5c2bf149c2054 100644 (file)
--- a/gcc/c/c-typeck.c
+++ b/gcc/c/c-typeck.c
@@ -13689,6 +13689,7 @@ handle_omp_array_sections (tree c, enum c_omp_region_type ort)
        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);
       if (OMP_CLAUSE_MAP_KIND (c2) != GOMP_MAP_FIRSTPRIVATE_POINTER
          && !c_mark_addressable (t))
        return false;
@@ -13959,7 +13960,8 @@ tree
 c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 {
   bitmap_head generic_head, firstprivate_head, lastprivate_head;
-  bitmap_head aligned_head, map_head, map_field_head, oacc_reduction_head;
+  bitmap_head aligned_head, map_head, map_field_head, map_firstprivate_head;
+  bitmap_head oacc_reduction_head;
   tree c, t, type, *pc;
   tree simdlen = NULL_TREE, safelen = NULL_TREE;
   bool branch_seen = false;
@@ -13980,7 +13982,7 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
      has been seen, -2 if mixed inscan/normal reduction diagnosed.  */
   int reduction_seen = 0;
   bool allocate_seen = false;
-  bool firstprivate_implicit_moved = false;
+  bool implicit_moved = false;
   bool oacc_gang_seen = false;
 
   bitmap_obstack_initialize (NULL);
@@ -13991,6 +13993,7 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
   /* If ort == C_ORT_OMP_DECLARE_SIMD used as uniform_head instead.  */
   bitmap_initialize (&map_head, &bitmap_default_obstack);
   bitmap_initialize (&map_field_head, &bitmap_default_obstack);
+  bitmap_initialize (&map_firstprivate_head, &bitmap_default_obstack);
   /* If ort == C_ORT_OMP used as nontemporal_head or use_device_xxx_head
      instead.  */
   bitmap_initialize (&oacc_reduction_head, &bitmap_default_obstack);
@@ -14438,17 +14441,25 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
          break;
 
        case OMP_CLAUSE_FIRSTPRIVATE:
-         if (OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT (c)
-             && !firstprivate_implicit_moved)
+         if (OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT (c) && !implicit_moved)
            {
-             firstprivate_implicit_moved = true;
-             /* Move firstprivate clauses with
-                OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT set to the end of
+           move_implicit:
+             implicit_moved = true;
+             /* Move firstprivate and map clauses with
+                OMP_CLAUSE_{FIRSTPRIVATE,MAP}_IMPLICIT set to the end of
                 clauses chain.  */
-             tree cl = NULL, *pc1 = pc, *pc2 = &cl;
+             tree cl1 = NULL_TREE, cl2 = NULL_TREE;
+             tree *pc1 = pc, *pc2 = &cl1, *pc3 = &cl2;
              while (*pc1)
                if (OMP_CLAUSE_CODE (*pc1) == OMP_CLAUSE_FIRSTPRIVATE
                    && OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT (*pc1))
+                 {
+                   *pc3 = *pc1;
+                   pc3 = &OMP_CLAUSE_CHAIN (*pc3);
+                   *pc1 = OMP_CLAUSE_CHAIN (*pc1);
+                 }
+               else if (OMP_CLAUSE_CODE (*pc1) == OMP_CLAUSE_MAP
+                        && OMP_CLAUSE_MAP_IMPLICIT (*pc1))
                  {
                    *pc2 = *pc1;
                    pc2 = &OMP_CLAUSE_CHAIN (*pc2);
@@ -14456,10 +14467,10 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                  }
                else
                  pc1 = &OMP_CLAUSE_CHAIN (*pc1);
-             *pc2 = NULL;
-             *pc1 = cl;
-             if (pc1 != pc)
-               continue;
+             *pc3 = NULL;
+             *pc2 = cl2;
+             *pc1 = cl1;
+             continue;
            }
          t = OMP_CLAUSE_DECL (c);
          need_complete = true;
@@ -14470,6 +14481,10 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                        "%qE is not a variable in clause %<firstprivate%>", t);
              remove = true;
            }
+         else if (OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT (c)
+                  && !OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT_TARGET (c)
+                  && bitmap_bit_p (&map_firstprivate_head, DECL_UID (t)))
+           remove = true;
          else if (bitmap_bit_p (&generic_head, DECL_UID (t))
                   || bitmap_bit_p (&firstprivate_head, DECL_UID (t)))
            {
@@ -14722,6 +14737,9 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
          break;
 
        case OMP_CLAUSE_MAP:
+         if (OMP_CLAUSE_MAP_IMPLICIT (c) && !implicit_moved)
+           goto move_implicit;
+         /* FALLTHRU */
        case OMP_CLAUSE_TO:
        case OMP_CLAUSE_FROM:
        case OMP_CLAUSE__CACHE_:
@@ -14760,6 +14778,16 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                          t = TREE_OPERAND (t, 0);
                          STRIP_NOPS (t);
                        }
+                     if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP
+                         && OMP_CLAUSE_MAP_IMPLICIT (c)
+                         && (bitmap_bit_p (&map_head, DECL_UID (t))
+                             || bitmap_bit_p (&map_field_head, DECL_UID (t))
+                             || bitmap_bit_p (&map_firstprivate_head,
+                                              DECL_UID (t))))
+                       {
+                         remove = true;
+                         break;
+                       }
                      if (bitmap_bit_p (&map_field_head, DECL_UID (t)))
                        break;
                      if (bitmap_bit_p (&map_head, DECL_UID (t)))
@@ -14923,6 +14951,12 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                        omp_clause_code_name[OMP_CLAUSE_CODE (c)]);
              remove = true;
            }
+         else if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP
+                  && OMP_CLAUSE_MAP_IMPLICIT (c)
+                  && (bitmap_bit_p (&map_head, DECL_UID (t))
+                      || bitmap_bit_p (&map_field_head, DECL_UID (t))
+                      || bitmap_bit_p (&map_firstprivate_head, DECL_UID (t))))
+           remove = true;
          else if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP
                   && OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_FIRSTPRIVATE_POINTER)
            {
@@ -14944,7 +14978,10 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                  remove = true;
                }
              else
-               bitmap_set_bit (&generic_head, DECL_UID (t));
+               {
+                 bitmap_set_bit (&generic_head, DECL_UID (t));
+                 bitmap_set_bit (&map_firstprivate_head, DECL_UID (t));
+               }
            }
          else if (bitmap_bit_p (&map_head, DECL_UID (t))
                   && !bitmap_bit_p (&map_field_head, DECL_UID (t))

diff --git a/gcc/cp/ChangeLog.omp b/gcc/cp/ChangeLog.omp
index d4d2d421f71fae82a0b2a7fe7af093f0f999596f..5e2685b3e60f3991f5828edb6963547fef490612 100644 (file)
--- a/gcc/cp/ChangeLog.omp
+++ b/gcc/cp/ChangeLog.omp
@@ -1,3 +1,18 @@
+2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
+
+       Backported from master:
+       2021-05-28  Jakub Jelinek  <jakub@redhat.com>
+
+       * semantics.c (handle_omp_array_sections): Copy
+       OMP_CLAUSE_MAP_IMPLICIT.
+       (finish_omp_clauses): Move not just OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT
+       marked clauses last, but also OMP_CLAUSE_MAP_IMPLICIT.  Add
+       map_firstprivate_head bitmap, set it for GOMP_MAP_FIRSTPRIVATE_POINTER
+       maps and silently remove OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT if it is
+       present too.  For OMP_CLAUSE_MAP_IMPLICIT silently remove the clause
+       if present in map_head, map_field_head or map_firstprivate_head
+       bitmaps.
+
 2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
 
        Backported from master:

diff --git a/gcc/cp/semantics.c b/gcc/cp/semantics.c
index 71d99e6f1e5d48c802081d40e9bac0da6bcfa83d..66b44ddb5903baf3394768822051229a070de1d7 100644 (file)
--- a/gcc/cp/semantics.c
+++ b/gcc/cp/semantics.c
@@ -5596,6 +5596,7 @@ handle_omp_array_sections (tree c, enum c_omp_region_type ort)
            }
          else
            OMP_CLAUSE_SET_MAP_KIND (c2, GOMP_MAP_FIRSTPRIVATE_POINTER);
+         OMP_CLAUSE_MAP_IMPLICIT (c2) = OMP_CLAUSE_MAP_IMPLICIT (c);
          if (OMP_CLAUSE_MAP_KIND (c2) != GOMP_MAP_FIRSTPRIVATE_POINTER
              && !cxx_mark_addressable (t))
            return false;
@@ -5622,6 +5623,7 @@ handle_omp_array_sections (tree c, enum c_omp_region_type ort)
              tree c3 = build_omp_clause (OMP_CLAUSE_LOCATION (c),
                                          OMP_CLAUSE_MAP);
              OMP_CLAUSE_SET_MAP_KIND (c3, OMP_CLAUSE_MAP_KIND (c2));
+             OMP_CLAUSE_MAP_IMPLICIT (c2) = OMP_CLAUSE_MAP_IMPLICIT (c);
              OMP_CLAUSE_DECL (c3) = ptr;
              if (OMP_CLAUSE_MAP_KIND (c2) == GOMP_MAP_ALWAYS_POINTER
                  || OMP_CLAUSE_MAP_KIND (c2) == GOMP_MAP_ATTACH_DETACH)
@@ -6561,7 +6563,8 @@ tree
 finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 {
   bitmap_head generic_head, firstprivate_head, lastprivate_head;
-  bitmap_head aligned_head, map_head, map_field_head, oacc_reduction_head;
+  bitmap_head aligned_head, map_head, map_field_head, map_firstprivate_head;
+  bitmap_head oacc_reduction_head;
   tree c, t, *pc;
   tree safelen = NULL_TREE;
   bool branch_seen = false;
@@ -6579,7 +6582,7 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
   bool allocate_seen = false;
   tree detach_seen = NULL_TREE;
   bool mergeable_seen = false;
-  bool firstprivate_implicit_moved = false;
+  bool implicit_moved = false;
   bool oacc_gang_seen = false;
 
   bitmap_obstack_initialize (NULL);
@@ -6590,6 +6593,7 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
   /* If ort == C_ORT_OMP_DECLARE_SIMD used as uniform_head instead.  */
   bitmap_initialize (&map_head, &bitmap_default_obstack);
   bitmap_initialize (&map_field_head, &bitmap_default_obstack);
+  bitmap_initialize (&map_firstprivate_head, &bitmap_default_obstack);
   /* If ort == C_ORT_OMP used as nontemporal_head or use_device_xxx_head
      instead.  */
   bitmap_initialize (&oacc_reduction_head, &bitmap_default_obstack);
@@ -6917,17 +6921,25 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
          break;
 
        case OMP_CLAUSE_FIRSTPRIVATE:
-         if (OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT (c)
-             && !firstprivate_implicit_moved)
+         if (OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT (c) && !implicit_moved)
            {
-             firstprivate_implicit_moved = true;
-             /* Move firstprivate clauses with
-                OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT set to the end of
+           move_implicit:
+             implicit_moved = true;
+             /* Move firstprivate and map clauses with
+                OMP_CLAUSE_{FIRSTPRIVATE,MAP}_IMPLICIT set to the end of
                 clauses chain.  */
-             tree cl = NULL, *pc1 = pc, *pc2 = &cl;
+             tree cl1 = NULL_TREE, cl2 = NULL_TREE;
+             tree *pc1 = pc, *pc2 = &cl1, *pc3 = &cl2;
              while (*pc1)
                if (OMP_CLAUSE_CODE (*pc1) == OMP_CLAUSE_FIRSTPRIVATE
                    && OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT (*pc1))
+                 {
+                   *pc3 = *pc1;
+                   pc3 = &OMP_CLAUSE_CHAIN (*pc3);
+                   *pc1 = OMP_CLAUSE_CHAIN (*pc1);
+                 }
+               else if (OMP_CLAUSE_CODE (*pc1) == OMP_CLAUSE_MAP
+                        && OMP_CLAUSE_MAP_IMPLICIT (*pc1))
                  {
                    *pc2 = *pc1;
                    pc2 = &OMP_CLAUSE_CHAIN (*pc2);
@@ -6935,10 +6947,10 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                  }
                else
                  pc1 = &OMP_CLAUSE_CHAIN (*pc1);
-             *pc2 = NULL;
-             *pc1 = cl;
-             if (pc1 != pc)
-               continue;
+             *pc3 = NULL;
+             *pc2 = cl2;
+             *pc1 = cl1;
+             continue;
            }
          t = omp_clause_decl_field (OMP_CLAUSE_DECL (c));
          if (t)
@@ -6969,6 +6981,10 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                          t);
              remove = true;
            }
+         else if (OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT (c)
+                  && !OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT_TARGET (c)
+                  && bitmap_bit_p (&map_firstprivate_head, DECL_UID (t)))
+           remove = true;
          else if (bitmap_bit_p (&generic_head, DECL_UID (t))
                   || bitmap_bit_p (&firstprivate_head, DECL_UID (t)))
            {
@@ -7685,6 +7701,9 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
          break;
 
        case OMP_CLAUSE_MAP:
+         if (OMP_CLAUSE_MAP_IMPLICIT (c) && !implicit_moved)
+           goto move_implicit;
+         /* FALLTHRU */
        case OMP_CLAUSE_TO:
        case OMP_CLAUSE_FROM:
        case OMP_CLAUSE__CACHE_:
@@ -7721,6 +7740,16 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                          t = TREE_OPERAND (t, 0);
                          STRIP_NOPS (t);
                        }
+                     if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP
+                         && OMP_CLAUSE_MAP_IMPLICIT (c)
+                         && (bitmap_bit_p (&map_head, DECL_UID (t))
+                             || bitmap_bit_p (&map_field_head, DECL_UID (t))
+                             || bitmap_bit_p (&map_firstprivate_head,
+                                              DECL_UID (t))))
+                       {
+                         remove = true;
+                         break;
+                       }
                      if (bitmap_bit_p (&map_field_head, DECL_UID (t)))
                        break;
                      if (bitmap_bit_p (&map_head, DECL_UID (t)))
@@ -7917,6 +7946,13 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                        "%qD is not a pointer variable", t);
              remove = true;
            }
+         else if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP
+                  && OMP_CLAUSE_MAP_IMPLICIT (c)
+                  && (bitmap_bit_p (&map_head, DECL_UID (t))
+                      || bitmap_bit_p (&map_field_head, DECL_UID (t))
+                      || bitmap_bit_p (&map_firstprivate_head,
+                                       DECL_UID (t))))
+           remove = true;
          else if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP
                   && OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_FIRSTPRIVATE_POINTER)
            {
@@ -7938,7 +7974,10 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
                  remove = true;
                }
              else
-               bitmap_set_bit (&generic_head, DECL_UID (t));
+               {
+                 bitmap_set_bit (&generic_head, DECL_UID (t));
+                 bitmap_set_bit (&map_firstprivate_head, DECL_UID (t));
+               }
            }
          else if (bitmap_bit_p (&map_head, DECL_UID (t))
                   && !bitmap_bit_p (&map_field_head, DECL_UID (t))

diff --git a/gcc/testsuite/ChangeLog.omp b/gcc/testsuite/ChangeLog.omp
index 334bc07d2db2c80d13c5c222d262a4d13980a404..f2b44536d36f1b1855237aca16cd6c96097a2b45 100644 (file)
--- a/gcc/testsuite/ChangeLog.omp
+++ b/gcc/testsuite/ChangeLog.omp
@@ -1,3 +1,13 @@
+2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
+
+       Backported from master:
+       2021-05-28  Jakub Jelinek  <jakub@redhat.com>
+
+       * c-c++-common/gomp/pr99928-8.c: Remove all xfails.
+       * c-c++-common/gomp/pr99928-9.c: Likewise.
+       * c-c++-common/gomp/pr99928-10.c: Likewise.
+       * c-c++-common/gomp/pr99928-16.c: New test.
+
 2021-05-28  Tobias Burnus  <tobias@codesourcery.com>
 
        Backported from master:

diff --git a/gcc/testsuite/c-c++-common/gomp/pr99928-10.c b/gcc/testsuite/c-c++-common/gomp/pr99928-10.c
index d2987a1d9b46915494f1543466c0ad3972019449..6c44600b07e42a0e0a68641d31ac6e5df1f6fa81 100644 (file)
--- a/gcc/testsuite/c-c++-common/gomp/pr99928-10.c
+++ b/gcc/testsuite/c-c++-common/gomp/pr99928-10.c
@@ -95,22 +95,22 @@ bar (void)
     #pragma omp section
     r12[1]++;
   }
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 60\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r13 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 60\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r13 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r13\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r13 \\+ 4" "gimple" } } */
   #pragma omp target parallel reduction(+:r13[1:15])
   r13[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 64\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r14 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 64\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r14 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r14" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r14 \\+ 4" "gimple" } } *//* FIXME: This should be on for instead.  */
   /* { dg-final { scan-tree-dump-not "omp for\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r14 \\+ 4" "gimple" } } *//* FIXME.  */
   #pragma omp target parallel for reduction(+:r14[1:16])
   for (int i = 0; i < 64; i++)
     r14[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 68\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r15 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 68\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r15 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r15\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r15 \\+ 4" "gimple" } } *//* FIXME: This should be on for instead.  */
   /* { dg-final { scan-tree-dump-not "omp for\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r15 \\+ 4" "gimple" } } *//* FIXME.  */
@@ -118,31 +118,31 @@ bar (void)
   #pragma omp target parallel for simd reduction(+:r15[1:17])
   for (int i = 0; i < 64; i++)
     r15[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 72\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r16 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r16\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 72\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r16 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r16\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*firstprivate\\(r16\\)" "gimple" } } *//* FIXME: Should be shared, but firstprivate is an optimization.  */
   /* { dg-final { scan-tree-dump "omp for\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r16 \\+ 4" "gimple" } } *//* NOTE: This is implementation detail.  */
   /* { dg-final { scan-tree-dump "omp simd\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r16 \\+ 4" "gimple" } } *//* NOTE: This is implementation detail.  */
   #pragma omp target parallel loop reduction(+:r16[1:18])
   for (int i = 0; i < 64; i++)
     r16[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 76\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r17 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 76\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r17 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r17\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r17 \\+ 4" "gimple" } } */
   #pragma omp target teams reduction(+:r17[1:19])
   r17[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 80\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r18 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 80\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r18 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r18\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r18 \\+ 4" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r18 \\+ 4" "gimple" } } */
   #pragma omp target teams distribute reduction(+:r18[1:20])
   for (int i = 0; i < 64; i++)
     r18[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 84\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r19 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 84\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r19 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r19\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r19 \\+ 4" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r19 \\+ 4" "gimple" } } */
@@ -151,8 +151,8 @@ bar (void)
   #pragma omp target teams distribute parallel for reduction(+:r19[1:21])
   for (int i = 0; i < 64; i++)
     r19[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 88\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r20 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 88\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r20 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r20\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r20 \\+ 4" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r20 \\+ 4" "gimple" } } */
@@ -162,8 +162,8 @@ bar (void)
   #pragma omp target teams distribute parallel for simd reduction(+:r20[1:22])
   for (int i = 0; i < 64; i++)
     r20[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 92\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r21 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 92\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r21 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r21\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r21 \\+ 4" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r21 \\+ 4" "gimple" } } */
@@ -171,9 +171,9 @@ bar (void)
   #pragma omp target teams distribute simd reduction(+:r21[1:23])
   for (int i = 0; i < 64; i++)
     r21[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 96\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r22 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r22\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 96\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r22 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r22\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*firstprivate\\(r22\\)" "gimple" } } *//* FIXME: Should be shared, but firstprivate is an optimization.  */
   /* { dg-final { scan-tree-dump "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r22 \\+ 4" "gimple" } } *//* NOTE: This is implementation detail.  */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*firstprivate\\(r22\\)" "gimple" } } *//* NOTE: This is implementation detail.  */
@@ -182,8 +182,8 @@ bar (void)
   #pragma omp target teams loop reduction(+:r22[1:24])
   for (int i = 0; i < 64; i++)
     r22[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 100\\\]" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r23 \\\[pointer assign, bias: 4\\\]\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:\\\*_\[0-9]* \\\[len: 100\\\]" "gimple" } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(firstprivate:r23 \\\[pointer assign, bias: 4\\\]\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r23\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp simd\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*\\)r23 \\+ 4" "gimple" } } */
   #pragma omp target simd reduction(+:r23[1:25])

diff --git a/gcc/testsuite/c-c++-common/gomp/pr99928-16.c b/gcc/testsuite/c-c++-common/gomp/pr99928-16.c
new file mode 100644 (file)
index 0000000..84cd85d
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/pr99928-16.c
@@ -0,0 +1,16 @@
+/* PR middle-end/99928 */
+
+void
+foo (void)
+{
+  int a[6] = {};
+  #pragma omp target simd reduction(+:a[:3])
+  for (int i = 0; i < 6; i++)
+    a[0]++;
+  #pragma omp target simd reduction(+:a[:3]) map(always, tofrom: a)
+  for (int i = 0; i < 6; i++)
+    a[0]++;
+  #pragma omp target simd reduction(+:a[:3]) map(always, tofrom: a[:6])
+  for (int i = 0; i < 6; i++)
+    a[0]++;
+}

diff --git a/gcc/testsuite/c-c++-common/gomp/pr99928-8.c b/gcc/testsuite/c-c++-common/gomp/pr99928-8.c
index fc57573eaf5dd4c778ccbd97a9fa74b4a6e543ef..27e6ad1eb36d479b51f401b30540cc1b817fd32b 100644 (file)
--- a/gcc/testsuite/c-c++-common/gomp/pr99928-8.c
+++ b/gcc/testsuite/c-c++-common/gomp/pr99928-8.c
@@ -94,48 +94,48 @@ bar (void)
     #pragma omp section
     r12++;
   }
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r13" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r13\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r13" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r13\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:r13\\)" "gimple" } } */
   #pragma omp target parallel reduction(+:r13)
   r13++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r14" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r14\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r14" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r14\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:r14\\)" "gimple" } } *//* FIXME: This should be on for instead.  */
   /* { dg-final { scan-tree-dump-not "omp for\[^\n\r]*reduction\\(\\+:r14\\)" "gimple" } } *//* FIXME.  */
   #pragma omp target parallel for reduction(+:r14)
   for (int i = 0; i < 64; i++)
     r14++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r15" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r15\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r15" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r15\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:r15\\)" "gimple" } } *//* FIXME: This should be on for instead.  */
   /* { dg-final { scan-tree-dump-not "omp for\[^\n\r]*reduction\\(\\+:r15\\)" "gimple" } } *//* FIXME.  */
   /* { dg-final { scan-tree-dump "omp simd\[^\n\r]*reduction\\(\\+:r15\\)" "gimple" } } */
   #pragma omp target parallel for simd reduction(+:r15)
   for (int i = 0; i < 64; i++)
     r15++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r16" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r16\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r16" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r16\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*shared\\(r16\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp for\[^\n\r]*reduction\\(\\+:r16\\)" "gimple" } } *//* NOTE: This is implementation detail.  */
   /* { dg-final { scan-tree-dump "omp simd\[^\n\r]*reduction\\(\\+:r16\\)" "gimple" } } *//* NOTE: This is implementation detail.  */
   #pragma omp target parallel loop reduction(+:r16)
   for (int i = 0; i < 64; i++)
     r16++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r17" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r17\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r17" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r17\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:r17\\)" "gimple" } } */
   #pragma omp target teams reduction(+:r17)
   r17++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r18" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r18\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r18" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r18\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:r18\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:r18\\)" "gimple" } } */
   #pragma omp target teams distribute reduction(+:r18)
   for (int i = 0; i < 64; i++)
     r18++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r19" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r19\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r19" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r19\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:r19\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:r19\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:r19\\)" "gimple" } } *//* FIXME: This should be on for instead.  */
@@ -143,8 +143,8 @@ bar (void)
   #pragma omp target teams distribute parallel for reduction(+:r19)
   for (int i = 0; i < 64; i++)
     r19++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r20" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r20\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r20" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r20\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:r20\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:r20\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:r20\\)" "gimple" } } *//* FIXME: This should be on for instead.  */
@@ -153,16 +153,16 @@ bar (void)
   #pragma omp target teams distribute parallel for simd reduction(+:r20)
   for (int i = 0; i < 64; i++)
     r20++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r21" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r21\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r21" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r21\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:r21\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:r21\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp simd\[^\n\r]*reduction\\(\\+:r21\\)" "gimple" } } */
   #pragma omp target teams distribute simd reduction(+:r21)
   for (int i = 0; i < 64; i++)
     r21++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r22" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r22\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r22" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r22\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*shared\\(r22\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp distribute\[^\n\r]*reduction\\(\\+:r22\\)" "gimple" } } *//* NOTE: This is implementation detail.  */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*shared\\(r22\\)" "gimple" } } *//* NOTE: This is implementation detail.  */
@@ -171,8 +171,8 @@ bar (void)
   #pragma omp target teams loop reduction(+:r22)
   for (int i = 0; i < 64; i++)
     r22++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r23" "gimple" { xfail *-*-* } } } */
-  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r23\\)" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r23" "gimple" } } */
+  /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r23\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp simd\[^\n\r]*reduction\\(\\+:r23\\)" "gimple" } } */
   #pragma omp target simd reduction(+:r23)
   for (int i = 0; i < 64; i++)

diff --git a/gcc/testsuite/c-c++-common/gomp/pr99928-9.c b/gcc/testsuite/c-c++-common/gomp/pr99928-9.c
index c049bb0d2575a8c25a0addeb9d21171f961d3a22..862352776674e412c913d745ea08e0222f89a177 100644 (file)
--- a/gcc/testsuite/c-c++-common/gomp/pr99928-9.c
+++ b/gcc/testsuite/c-c++-common/gomp/pr99928-9.c
@@ -94,19 +94,19 @@ bar (void)
     #pragma omp section
     r12[1]++;
   }
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r13\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r13\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r13\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r13 \\+ 4" "gimple" } } */
   #pragma omp target parallel reduction(+:r13[1:2])
   r13[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r14\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r14\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r14" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r14 \\+ 4" "gimple" } } *//* FIXME: This should be on for instead.  */
   /* { dg-final { scan-tree-dump-not "omp for\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r14 \\+ 4" "gimple" } } *//* FIXME.  */
   #pragma omp target parallel for reduction(+:r14[1:2])
   for (int i = 0; i < 64; i++)
     r14[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r15\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r15\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r15\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r15 \\+ 4" "gimple" } } *//* FIXME: This should be on for instead.  */
   /* { dg-final { scan-tree-dump-not "omp for\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r15 \\+ 4" "gimple" } } *//* FIXME.  */
@@ -114,7 +114,7 @@ bar (void)
   #pragma omp target parallel for simd reduction(+:r15[1:2])
   for (int i = 0; i < 64; i++)
     r15[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r16\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r16\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r16\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp parallel\[^\n\r]*shared\\(r16\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp for\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r16 \\+ 4" "gimple" } } *//* NOTE: This is implementation detail.  */
@@ -122,19 +122,19 @@ bar (void)
   #pragma omp target parallel loop reduction(+:r16[1:2])
   for (int i = 0; i < 64; i++)
     r16[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r17\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r17\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r17\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r17 \\+ 4" "gimple" } } */
   #pragma omp target teams reduction(+:r17[1:2])
   r17[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r18\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r18\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r18\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r18 \\+ 4" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r18 \\+ 4" "gimple" } } */
   #pragma omp target teams distribute reduction(+:r18[1:2])
   for (int i = 0; i < 64; i++)
     r18[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r19\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r19\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r19\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r19 \\+ 4" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r19 \\+ 4" "gimple" } } */
@@ -143,7 +143,7 @@ bar (void)
   #pragma omp target teams distribute parallel for reduction(+:r19[1:2])
   for (int i = 0; i < 64; i++)
     r19[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r20\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r20\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r20\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r20 \\+ 4" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r20 \\+ 4" "gimple" } } */
@@ -153,7 +153,7 @@ bar (void)
   #pragma omp target teams distribute parallel for simd reduction(+:r20[1:2])
   for (int i = 0; i < 64; i++)
     r20[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r21\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r21\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r21\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r21 \\+ 4" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r21 \\+ 4" "gimple" } } */
@@ -161,7 +161,7 @@ bar (void)
   #pragma omp target teams distribute simd reduction(+:r21[1:2])
   for (int i = 0; i < 64; i++)
     r21[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r22\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r22\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r22\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp teams\[^\n\r]*shared\\(r22\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp distribute\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r22 \\+ 4" "gimple" } } *//* NOTE: This is implementation detail.  */
@@ -171,7 +171,7 @@ bar (void)
   #pragma omp target teams loop reduction(+:r22[1:2])
   for (int i = 0; i < 64; i++)
     r22[1]++;
-  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r23\\\[1\\\] \\\[len: 8\\\]" "gimple" { xfail *-*-* } } } */
+  /* { dg-final { scan-tree-dump "omp target\[^\n\r]*map\\(tofrom:r23\\\[1\\\] \\\[len: 8\\\]" "gimple" } } */
   /* { dg-final { scan-tree-dump-not "omp target\[^\n\r]*firstprivate\\(r23\\)" "gimple" } } */
   /* { dg-final { scan-tree-dump "omp simd\[^\n\r]*reduction\\(\\+:MEM\[^\n\r]*&r23 \\+ 4" "gimple" } } */
   #pragma omp target simd reduction(+:r23[1:2])

diff --git a/gcc/tree.h b/gcc/tree.h
index 5817a0ff3b23e5d0d79ac937cd3fdcc5f3ac9da5..ec01f6593edc8c7e572914b1cd25355deeb4b4b5 100644 (file)
--- a/gcc/tree.h
+++ b/gcc/tree.h
@@ -1660,6 +1660,11 @@ class auto_suppress_location_wrappers
    variable.  */
 #define OMP_CLAUSE_MAP_IN_REDUCTION(NODE) \
   TREE_PRIVATE (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_MAP))
+/* Nonzero on map clauses added implicitly for reduction clauses on combined
+   or composite constructs.  They shall be removed if there is an explicit
+   map clause.  */
+#define OMP_CLAUSE_MAP_IMPLICIT(NODE) \
+  (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_MAP)->base.default_def_flag)
 /* Nonzero if this map clause was created through implicit data-mapping
    rules. */
 #define OMP_CLAUSE_MAP_IMPLICIT_P(NODE) \