summaryrefslogtreecommitdiff
diff options
context:
space:
mode:
authorKwok Cheung Yeung <kcyeung@baylibre.com>2025-04-16 11:43:00 +0100
committerKwok Cheung Yeung <kcyeung@baylibre.com>2025-04-17 23:29:54 +0100
commit89e6586616e84564902b4721364a091daa6cd0a3 (patch)
tree4346ea7f12a67406bd87e6fa5db848bdcd893a17
parent2afb6281a5a8711b82aa744e71c181bab6537e3c (diff)
openmp, fortran: Revert to using tree expressions when translating Fortran OpenMP array sections
In the patch 'OpenACC 2.7: Implement reductions for arrays and records', temporaries are used to hold the decl and bias of clauses resulting from array sections, which breaks some assumptions made for map iterator support. This patch reverts the change for OpenMP only. gcc/fortran/ * trans-openmp.cc (gfc_trans_omp_array_section): Use temporaries only when translating OpenACC. gcc/testsuite/ * gfortran.dg/gomp/target-enter-exit-data.f90: Revert expected tree dumps.
-rw-r--r--gcc/fortran/ChangeLog.omp5
-rw-r--r--gcc/fortran/trans-openmp.cc24
-rw-r--r--gcc/testsuite/ChangeLog.omp5
-rw-r--r--gcc/testsuite/gfortran.dg/gomp/target-enter-exit-data.f908
4 files changed, 30 insertions, 12 deletions
diff --git a/gcc/fortran/ChangeLog.omp b/gcc/fortran/ChangeLog.omp
index 2b9d095cfa3..b0846c29029 100644
--- a/gcc/fortran/ChangeLog.omp
+++ b/gcc/fortran/ChangeLog.omp
@@ -1,5 +1,10 @@
2025-04-17 Kwok Cheung Yeung <kcyeung@baylibre.com>
+ * trans-openmp.cc (gfc_trans_omp_array_section): Use temporaries only
+ when translating OpenACC.
+
+2025-04-17 Kwok Cheung Yeung <kcyeung@baylibre.com>
+
* gfortran.h (struct gfc_omp_namelist): Move udm field into a new
union u3.
* match.cc (gfc_free_omp_namelist): Change references to u2.udm to
diff --git a/gcc/fortran/trans-openmp.cc b/gcc/fortran/trans-openmp.cc
index 7c76360a56e..1babad8aa74 100644
--- a/gcc/fortran/trans-openmp.cc
+++ b/gcc/fortran/trans-openmp.cc
@@ -4114,10 +4114,14 @@ gfc_trans_omp_array_section (stmtblock_t *block, toc_directive cd,
offset = build2 (TRUNC_DIV_EXPR, ptrdiff_type_node,
offset, fold_convert (ptrdiff_type_node, elemsz));
- tree offset_tmp = create_tmp_var (ptrdiff_type_node);
- gfc_add_expr_to_block (block, build2 (MODIFY_EXPR, ptrdiff_type_node,
- offset_tmp, offset));
- offset = offset_tmp;
+ if (!openmp)
+ {
+ tree offset_tmp = create_tmp_var (ptrdiff_type_node);
+ gfc_add_expr_to_block (block, build2 (MODIFY_EXPR,
+ ptrdiff_type_node,
+ offset_tmp, offset));
+ offset = offset_tmp;
+ }
offset = build4_loc (input_location, ARRAY_REF,
TREE_TYPE (TREE_TYPE (decl)),
decl, offset, NULL_TREE, NULL_TREE);
@@ -4137,12 +4141,16 @@ gfc_trans_omp_array_section (stmtblock_t *block, toc_directive cd,
OMP_CLAUSE_DECL (node3) = decl;
}
- tree ptr_tmp = create_tmp_var (ptrdiff_type_node);
ptr = fold_build2 (MINUS_EXPR, ptrdiff_type_node, ptr,
fold_convert (ptrdiff_type_node, ptr2));
- gfc_add_expr_to_block (block, build2 (MODIFY_EXPR, ptrdiff_type_node,
- ptr_tmp, ptr));
- OMP_CLAUSE_SIZE (node3) = ptr_tmp;
+ if (!openmp)
+ {
+ tree ptr_tmp = create_tmp_var (ptrdiff_type_node);
+ gfc_add_expr_to_block (block, build2 (MODIFY_EXPR, ptrdiff_type_node,
+ ptr_tmp, ptr));
+ ptr = ptr_tmp;
+ }
+ OMP_CLAUSE_SIZE (node3) = ptr;
if (n->u.map.readonly)
OMP_CLAUSE_MAP_POINTS_TO_READONLY (node3) = 1;
diff --git a/gcc/testsuite/ChangeLog.omp b/gcc/testsuite/ChangeLog.omp
index 55b0ec12dcb..331cb11af96 100644
--- a/gcc/testsuite/ChangeLog.omp
+++ b/gcc/testsuite/ChangeLog.omp
@@ -1,5 +1,10 @@
2025-04-17 Kwok Cheung Yeung <kcyeung@baylibre.com>
+ * gfortran.dg/gomp/target-enter-exit-data.f90: Revert expected tree
+ dumps.
+
+2025-04-17 Kwok Cheung Yeung <kcyeung@baylibre.com>
+
* c-c++-common/gomp/target-update-iterators-1.c: New.
* c-c++-common/gomp/target-update-iterators-2.c: New.
* c-c++-common/gomp/target-update-iterators-3.c: New.
diff --git a/gcc/testsuite/gfortran.dg/gomp/target-enter-exit-data.f90 b/gcc/testsuite/gfortran.dg/gomp/target-enter-exit-data.f90
index dc267c2af2e..74eb894c04c 100644
--- a/gcc/testsuite/gfortran.dg/gomp/target-enter-exit-data.f90
+++ b/gcc/testsuite/gfortran.dg/gomp/target-enter-exit-data.f90
@@ -9,17 +9,17 @@ type(t) :: var
allocate (var%arr(1:100))
!$omp target enter data map(to: var%arr(10:20))
-! { dg-final { scan-tree-dump-times {(?n)#pragma omp target enter data map\(to:\*\(integer\(kind=[0-9]+\)\[0:\] \*\) parm\.[0-9]+\.data \[len: D\.[0-9]+ \* [0-9]+\]\) map\(to:var\.arr \[pointer set, len: [0-9]+\]\) map\(attach_detach:\(integer\(kind=[0-9]+\)\[0:\] \*\) var\.arr\.data \[bias: D\.[0-9]+\]\)$} 1 "original" } }
+! { dg-final { scan-tree-dump-times {(?n)#pragma omp target enter data map\(to:\*\(integer\(kind=[0-9]+\)\[0:\] \*\) parm\.[0-9]+\.data \[len: D\.[0-9]+ \* [0-9]+\]\) map\(to:var\.arr \[pointer set, len: [0-9]+\]\) map\(attach_detach:\(integer\(kind=[0-9]+\)\[0:\] \*\) var\.arr\.data \[bias: \(integer\(kind=[0-9]+\)\) parm\.[0-9]+\.data - \(integer\(kind=[0-9]+\)\) var\.arr\.data\]\)$} 1 "original" } }
!$omp target exit data map(release: var%arr(10:20))
-! { dg-final { scan-tree-dump-times {(?n)#pragma omp target exit data map\(release:\*\(integer\(kind=[0-9]+\)\[0:\] \*\) parm\.[0-9]+\.data \[len: D\.[0-9]+ \* [0-9]+\]\) map\(release:var\.arr \[pointer set, len: [0-9]+\]\) map\(attach_detach:\(integer\(kind=[0-9]+\)\[0:\] \*\) var\.arr\.data \[bias: D\.[0-9]+\]\)$} 1 "original" } }
+! { dg-final { scan-tree-dump-times {(?n)#pragma omp target exit data map\(release:\*\(integer\(kind=[0-9]+\)\[0:\] \*\) parm\.[0-9]+\.data \[len: D\.[0-9]+ \* [0-9]+\]\) map\(release:var\.arr \[pointer set, len: [0-9]+\]\) map\(attach_detach:\(integer\(kind=[0-9]+\)\[0:\] \*\) var\.arr\.data \[bias: \(integer\(kind=[0-9]+\)\) parm\.[0-9]+\.data - \(integer\(kind=[0-9]+\)\) var\.arr\.data\]\)$} 1 "original" } }
!$omp target enter data map(alloc: var%arr(20:30))
-! { dg-final { scan-tree-dump-times {(?n)#pragma omp target enter data map\(alloc:\*\(integer\(kind=[0-9]+\)\[0:\] \*\) parm\.[0-9]+\.data \[len: D\.[0-9]+ \* [0-9]+\]\) map\(to:var\.arr \[pointer set, len: [0-9]+\]\) map\(attach_detach:\(integer\(kind=[0-9]+\)\[0:\] \*\) var\.arr\.data \[bias: D\.[0-9]+\]\)$} 1 "original" } }
+! { dg-final { scan-tree-dump-times {(?n)#pragma omp target enter data map\(alloc:\*\(integer\(kind=[0-9]+\)\[0:\] \*\) parm\.[0-9]+\.data \[len: D\.[0-9]+ \* [0-9]+\]\) map\(to:var\.arr \[pointer set, len: [0-9]+\]\) map\(attach_detach:\(integer\(kind=[0-9]+\)\[0:\] \*\) var\.arr\.data \[bias: \(integer\(kind=[0-9]+\)\) parm\.[0-9]+\.data - \(integer\(kind=[0-9]+\)\) var\.arr\.data\]\)$} 1 "original" } }
!$omp target exit data map(delete: var%arr(20:30))
-! { dg-final { scan-tree-dump-times {(?n)#pragma omp target exit data map\(delete:\*\(integer\(kind=[0-9]+\)\[0:\] \*\) parm\.[0-9]+\.data \[len: D\.[0-9]+ \* [0-9]+\]\) map\(delete:var\.arr \[pointer set, len: [0-9]+\]\) map\(attach_detach:\(integer\(kind=[0-9]+\)\[0:\] \*\) var\.arr\.data \[bias: D\.[0-9]+\]\)$} 1 "original" } }
+! { dg-final { scan-tree-dump-times {(?n)#pragma omp target exit data map\(delete:\*\(integer\(kind=[0-9]+\)\[0:\] \*\) parm\.[0-9]+\.data \[len: D\.[0-9]+ \* [0-9]+\]\) map\(delete:var\.arr \[pointer set, len: [0-9]+\]\) map\(attach_detach:\(integer\(kind=[0-9]+\)\[0:\] \*\) var\.arr\.data \[bias: \(integer\(kind=[0-9]+\)\) parm\.[0-9]+\.data - \(integer\(kind=[0-9]+\)\) var\.arr\.data\]\)$} 1 "original" } }
!$omp target enter data map(to: var%arr)