Hi,
This patch adds handling for the map-type-modifier 'close' in the map
clause in the Fortran parser (for C and C++ parsers the changes were
already committed).
'close' was introduced with OpenMP 5.0: "The close map-type-modifier is
a hint to the runtime to allocate memory close to the target device."
In OpenMP 5.0 'close' can be used beside/together with 'always' in a
list of map-type-modifiers.
This patch also considers the optional commas in the modifier list,
which the old code did not (although the comma after 'always' was
already optional in OpenMP 4.5).
Marcel
-----------------
Mentor Graphics (Deutschland) GmbH, Arnulfstrasse 201, 80634 München
Registergericht München HRB 106955, Geschäftsführer: Thomas Heurung, Frank
Thürauf
Fortran/OpenMP: Add support for 'close' in map clause
gcc/fortran/ChangeLog:
* openmp.c (gfc_match_omp_clauses): Support map-type-modifier 'close'.
gcc/testsuite/ChangeLog:
* gfortran.dg/gomp/map-6.f90: New test.
* gfortran.dg/gomp/map-7.f90: New test.
* gfortran.dg/gomp/map-8.f90: New test.
diff --git a/gcc/fortran/openmp.c b/gcc/fortran/openmp.c
index 7eeabff..bec852a 100644
--- a/gcc/fortran/openmp.c
+++ b/gcc/fortran/openmp.c
@@ -1710,10 +1710,21 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const
omp_mask mask,
&& gfc_match ("map ( ") == MATCH_YES)
{
locus old_loc2 = gfc_current_locus;
- bool always = false;
+
+ int always = 0;
+ int close = 0;
+ for (;;)
+ {
+ if (gfc_match ("always ") == MATCH_YES)
+ always++;
+ else if (gfc_match ("close ") == MATCH_YES)
+ close++;
+ else
+ break;
+ gfc_match (", ");
+ }
+
gfc_omp_map_op map_op = OMP_MAP_TOFROM;
- if (gfc_match ("always , ") == MATCH_YES)
- always = true;
if (gfc_match ("alloc : ") == MATCH_YES)
map_op = OMP_MAP_ALLOC;
else if (gfc_match ("tofrom : ") == MATCH_YES)
@@ -1726,11 +1737,24 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const
omp_mask mask,
map_op = OMP_MAP_RELEASE;
else if (gfc_match ("delete : ") == MATCH_YES)
map_op = OMP_MAP_DELETE;
- else if (always)
+ else
{
gfc_current_locus = old_loc2;
- always = false;
+ always = 0;
+ close = 0;
}
+
+ if (always > 1)
+ {
+ gfc_error ("too many %<always%> modifiers at %C");
+ break;
+ }
+ if (close > 1)
+ {
+ gfc_error ("too many %<close%> modifiers at %C");
+ break;
+ }
+
head = NULL;
if (gfc_match_omp_variable_list ("", &c->lists[OMP_LIST_MAP],
false, NULL, &head,
@@ -1741,8 +1765,8 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const
omp_mask mask,
n->u.map_op = map_op;
continue;
}
- else
- gfc_current_locus = old_loc;
+ gfc_current_locus = old_loc;
+ break;
}
if ((mask & OMP_CLAUSE_MERGEABLE) && !c->mergeable
&& gfc_match ("mergeable") == MATCH_YES)
diff --git a/gcc/testsuite/gfortran.dg/gomp/map-6.f90
b/gcc/testsuite/gfortran.dg/gomp/map-6.f90
new file mode 100644
index 0000000..309f845
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/map-6.f90
@@ -0,0 +1,50 @@
+! { dg-additional-options "-fdump-tree-original" }
+
+implicit none
+
+integer :: a, b, b1, b2, b3, b4, b5, b6
+
+!$omp target map(a)
+!$omp end target
+
+!$omp target map(to : a)
+!$omp end target
+
+!$omp target map(always to: a)
+!$omp end target
+!$omp target map(always, to: a)
+!$omp end target
+!$omp target map(close to: a)
+!$omp end target
+!$omp target map(close, to: a)
+!$omp end target
+
+!$omp target map(close always to:b1)
+!$omp end target
+!$omp target map(close, always to:b2)
+!$omp end target
+!$omp target map(close, always, to:b3)
+!$omp end target
+!$omp target map(always close to:b4)
+!$omp end target
+!$omp target map(always, close to:b5)
+!$omp end target
+!$omp target map(always, close, to:b6)
+!$omp end target
+
+
+!$omp target map (always to : a) map (close to : b)
+!$omp end target
+
+end
+
+! { dg-final { scan-tree-dump-not "map\\(\[^\n\r)]*close\[^\n\r)]*to:"
"original" } }
+
+! { dg-final { scan-tree-dump-times "#pragma omp target map\\(always,to:" 9
"original" } }
+
+! { dg-final { scan-tree-dump "#pragma omp target map\\(always,to:b1\\)"
"original" } }
+! { dg-final { scan-tree-dump "#pragma omp target map\\(always,to:b2\\)"
"original" } }
+! { dg-final { scan-tree-dump "#pragma omp target map\\(always,to:b3\\)"
"original" } }
+! { dg-final { scan-tree-dump "#pragma omp target map\\(always,to:b4\\)"
"original" } }
+! { dg-final { scan-tree-dump "#pragma omp target map\\(always,to:b5\\)"
"original" } }
+! { dg-final { scan-tree-dump "#pragma omp target map\\(always,to:b6\\)"
"original" } }
diff --git a/gcc/testsuite/gfortran.dg/gomp/map-7.f90
b/gcc/testsuite/gfortran.dg/gomp/map-7.f90
new file mode 100644
index 0000000..009c6d4
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/map-7.f90
@@ -0,0 +1,26 @@
+! { dg-additional-options "-fdump-tree-original" }
+
+implicit none
+
+integer :: a, b, close, always, to
+
+!$omp target map(close)
+!$omp end target
+
+!$omp target map(always)
+!$omp end target
+
+!$omp target map(always, close)
+!$omp end target
+
+!$omp target map(always, close, to : always, close, a)
+!$omp end target
+
+!$omp target map(to, always, close)
+!$omp end target
+
+end
+
+! { dg-final { scan-tree-dump-not "map\\(\[^\n\r)]*close\[^\n\r)]*to:"
"original" } }
+! { dg-final { scan-tree-dump "#pragma omp target map\\(always,to:always\\)
map\\(always,to:close\\) map\\(always,to:a\\)" "original" } }
+! { dg-final { scan-tree-dump-not "map\\(\[^\n\r)]*close\[^\n\r)]*to:"
"original" } }
diff --git a/gcc/testsuite/gfortran.dg/gomp/map-8.f90
b/gcc/testsuite/gfortran.dg/gomp/map-8.f90
new file mode 100644
index 0000000..92b802c
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/map-8.f90
@@ -0,0 +1,34 @@
+implicit none
+
+integer :: a
+
+!$omp target map(close, delete: a) ! { dg-error "TARGET with map-type other
than TO, FROM, TOFROM, or ALLOC on MAP clause at \\(1\\)" }
+!$omp end target
+
+!$omp target map(close) ! { dg-error "Symbol 'close' at \\(1\\) has no
IMPLICIT type" }
+!$omp end target
+
+!$omp target map(always) ! { dg-error "Symbol 'always' at \\(1\\) has no
IMPLICIT type" }
+!$omp end target
+
+!$omp target map(always, always, to : a) ! { dg-error "too many 'always'
modifiers" }
+! !$omp end target
+!$omp target map(always always, to : a) ! { dg-error "too many 'always'
modifiers" }
+! !$omp end target
+!$omp target map(always, always to : a) ! { dg-error "too many 'always'
modifiers" }
+! !$omp end target
+!$omp target map(always always to : a) ! { dg-error "too many 'always'
modifiers" }
+! !$omp end target
+!$omp target map(close, close, to : a) ! { dg-error "too many 'close'
modifiers" }
+! !$omp end target
+!$omp target map(close close, to : a) ! { dg-error "too many 'close'
modifiers" }
+! !$omp end target
+!$omp target map(close, close to : a) ! { dg-error "too many 'close'
modifiers" }
+! !$omp end target
+!$omp target map(close close to : a) ! { dg-error "too many 'close' modifiers"
}
+! !$omp end target
+
+!$omp target map(close close always always to : a) ! { dg-error "too many
'always' modifiers" }
+! !$omp end target
+
+end
