[gcc/devel/omp/gcc-16] openmp: Scale type precision of collapsed iterator variable

https://gcc.gnu.org/g:744de5d28df760c2decdc3fac9d55a9e7ac92707

commit 744de5d28df760c2decdc3fac9d55a9e7ac92707
Author: Kwok Cheung Yeung <[email protected]>
Date:   Mon Mar 1 14:15:30 2021 -0800

    openmp: Scale type precision of collapsed iterator variable
    
    This sets the type precision of the collapsed iterator variable to the
    sum of the precision of the collapsed loop variables, up to a maximum of
    sizeof(long long) (i.e. 64-bits).
    
    gcc/ChangeLog
            * omp-expand.cc (expand_oacc_for): Convert .tile variable to
            diff_type before multiplying.
            * omp-general.cc (omp_extract_for_data): Use accumulated precision
            of all collapsed for-loops as precision of iteration variable, up
            to the precision of a long long.
    
    libgomp/ChangeLog
            * testsuite/libgomp.c-c++-common/collapse-4.c: New.
            * testsuite/libgomp.fortran/collapse5.f90: New.

Diff:
---
 gcc/omp-expand.cc                                  |  5 ++-
 gcc/omp-general.cc                                 | 39 +++++++++++++++-------
 .../testsuite/libgomp.c-c++-common/collapse-4.c    | 23 +++++++++++++
 libgomp/testsuite/libgomp.fortran/collapse5.f90    | 23 +++++++++++++
 4 files changed, 77 insertions(+), 13 deletions(-)

diff --git a/gcc/omp-expand.cc b/gcc/omp-expand.cc
index 625dbf423d5f..d768e7215942 100644
--- a/gcc/omp-expand.cc
+++ b/gcc/omp-expand.cc
@@ -7856,7 +7856,10 @@ expand_oacc_for (struct omp_region *region, struct omp_for_data *fd)
       tile_size = create_tmp_var (diff_type, ".tile_size");
       expr = build_int_cst (diff_type, 1);
       for (int ix = 0; ix < fd->collapse; ix++)
-	expr = fold_build2 (MULT_EXPR, diff_type, counts[ix].tile, expr);
+	{
+	  tree tile = fold_convert (diff_type, counts[ix].tile);
+	  expr = fold_build2 (MULT_EXPR, diff_type, tile, expr);
+	}
       expr = force_gimple_operand_gsi (&gsi, expr, true,
 				       NULL_TREE, true, GSI_SAME_STMT);
       ass = gimple_build_assign (tile_size, expr);
diff --git a/gcc/omp-general.cc b/gcc/omp-general.cc
index fdd934f5223f..4fdf96541079 100644
--- a/gcc/omp-general.cc
+++ b/gcc/omp-general.cc
@@ -406,6 +406,7 @@ omp_extract_for_data (gomp_for *for_stmt, struct omp_for_data *fd,
 	  fd->non_rect = true;
 	}
     }
+  int accum_iter_precision = 0;
   for (i = 0; i < cnt; i++)
     {
       if (i == 0
@@ -489,19 +490,33 @@ omp_extract_for_data (gomp_for *for_stmt, struct omp_for_data *fd,
 	{
 	  if (fd->collapse == 1 && !fd->tiling)
 	    iter_type = TREE_TYPE (loop->v);
-	  else if (i == 0
-		   || TYPE_PRECISION (iter_type)
-		      < TYPE_PRECISION (TREE_TYPE (loop->v)))
+	  else
 	    {
-	      if (TREE_CODE (iter_type) == BITINT_TYPE
-		  || TREE_CODE (TREE_TYPE (loop->v)) == BITINT_TYPE)
-		iter_type
-		  = build_bitint_type (TYPE_PRECISION (TREE_TYPE (loop->v)),
-				       1);
-	      else
-		iter_type
-		  = build_nonstandard_integer_type
-			(TYPE_PRECISION (TREE_TYPE (loop->v)), 1);
+	      int loop_precision = TYPE_PRECISION (TREE_TYPE (loop->v));
+	      int iter_type_precision = 0;
+	      const int max_accum_precision
+		= TYPE_PRECISION (long_long_unsigned_type_node);
+
+	      accum_iter_precision += loop_precision;
+
+	      if (i == 0
+		  || (loop_precision >= max_accum_precision
+		      && loop_precision >= TYPE_PRECISION (iter_type)))
+		iter_type_precision = loop_precision;
+	      else if (TYPE_PRECISION (iter_type) < max_accum_precision)
+		iter_type_precision
+		  = MIN (1 << ceil_log2 (accum_iter_precision),
+			 max_accum_precision);
+
+	      if (iter_type_precision)
+		{
+		  if (TREE_CODE (iter_type) == BITINT_TYPE
+		      || TREE_CODE (TREE_TYPE (loop->v)) == BITINT_TYPE)
+		    iter_type = build_bitint_type (iter_type_precision, 1);
+		  else
+		    iter_type
+		      = build_nonstandard_integer_type (iter_type_precision, 1);
+		}
 	    }
 	}
       else if (iter_type != long_long_unsigned_type_node)
diff --git a/libgomp/testsuite/libgomp.c-c++-common/collapse-4.c b/libgomp/testsuite/libgomp.c-c++-common/collapse-4.c
new file mode 100644
index 000000000000..c0af29f54639
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/collapse-4.c
@@ -0,0 +1,23 @@
+/* { dg-do run } */
+
+#include <stdlib.h>
+
+int
+main (void)
+{
+  int i, j;
+  int count = 0;
+
+  #pragma omp parallel for collapse(2)
+    for (i = 0; i < 80000; i++)
+      for (j = 0; j < 80000; j++)
+	if (i == 66666 && j == 77777)
+	  /* In the collapsed loop space, this is iteration
+	     66666*80000+77777==5,333,357,777.  If the type of the iterator
+	     for the collapsed loop is only a 32-bit unsigned int, then this
+	     iteration will exceed its maximum range and be skipped.  */
+	  count++;
+
+  if (count != 1)
+    abort ();
+}
diff --git a/libgomp/testsuite/libgomp.fortran/collapse5.f90 b/libgomp/testsuite/libgomp.fortran/collapse5.f90
new file mode 100644
index 000000000000..5632d9bab020
--- /dev/null
+++ b/libgomp/testsuite/libgomp.fortran/collapse5.f90
@@ -0,0 +1,23 @@
+! { dg-do run }
+
+program collapse5
+  implicit none
+
+  integer :: i, j
+  integer :: count = 0
+
+  !$omp parallel do collapse (2)
+    do i = 1, 80000
+      do j = 1, 80000
+        if (i .eq. 66666 .and. j .eq. 77777) then
+	  ! In the collapsed loop space, this is iteration
+	  ! 66666*80000+77777==5,333,357,777.  If the type of the iterator
+	  ! for the collapsed loop is only a 32-bit unsigned int, then this
+	  ! iteration will exceed its maximum range and be skipped.
+	  count = count + 1
+	end if
+      end do
+    end do
+
+  if (count .ne. 1) stop 1
+end