From: Tobias Burnus Date: Fri, 20 Sep 2019 16:23:38 +0000 (+0200) Subject: Backport from mainline X-Git-Url: http://git.ipfire.org/cgi-bin/gitweb.cgi?a=commitdiff_plain;h=bb911c6f61333cf8e1018f303a66c35add30011c;p=thirdparty%2Fgcc.git Backport from mainline 2019-09-20 Tobias Burnus * openmp.c (gfc_resolve_oacc_declare): Reject all non variables but accept function result variables. * trans-openmp.c (gfc_trans_omp_clauses): Handle function-result variables for remaing cases. 2019-09-20 Tobias Burnus * gfortran.dg/goacc/parameter.f95: Change dg-error as it is now detected earlier. * gfortran.dg/goacc/pr85701.f90: Modify to use a separate result variable. * gfortran.dg/goacc/pr78260.f90: New. * gfortran.dg/goacc/pr78260-2.f90: New. * gfortran.dg/gomp/pr78260.f90: New. * gfortran.dg/gomp/pr78260-2.f90: New. * gfortran.dg/gomp/pr78260-3.f90: New. (cherry picked from openacc-gcc-9-branch commit 7d0d394c1c9bfb1b489c4ad3bd606d7a4765ae1d) --- diff --git a/gcc/fortran/ChangeLog.omp b/gcc/fortran/ChangeLog.omp index a54fb4e46144..87766c9c2a6b 100644 --- a/gcc/fortran/ChangeLog.omp +++ b/gcc/fortran/ChangeLog.omp @@ -1,3 +1,13 @@ +2019-09-20 Tobias Burnus + + Backported from mainline + 2019-09-20 Tobias Burnus + + * openmp.c (gfc_resolve_oacc_declare): Reject all + non variables but accept function result variables. + * trans-openmp.c (gfc_trans_omp_clauses): Handle + function-result variables for remaing cases. + 2019-09-17 Tobias Burnus * trans-expr.c (gfc_auto_dereference_var): Use passed loc argument. diff --git a/gcc/fortran/openmp.c b/gcc/fortran/openmp.c index fa0cd6f84a37..f8b4cb0aab16 100644 --- a/gcc/fortran/openmp.c +++ b/gcc/fortran/openmp.c @@ -6178,18 +6178,14 @@ gfc_resolve_oacc_declare (gfc_namespace *ns) for (n = oc->clauses->lists[list]; n; n = n->next) { n->sym->mark = 0; - if (n->sym->attr.function || n->sym->attr.subroutine) + if (n->sym->attr.flavor != FL_VARIABLE + && (n->sym->attr.flavor != FL_PROCEDURE + || n->sym->result != n->sym)) { gfc_error ("Object %qs is not a variable at %L", n->sym->name, &oc->loc); continue; } - if (n->sym->attr.flavor == FL_PARAMETER) - { - gfc_error ("PARAMETER object %qs is not allowed at %L", - n->sym->name, &oc->loc); - continue; - } if (n->expr && n->expr->ref->type == REF_ARRAY) { diff --git a/gcc/fortran/trans-openmp.c b/gcc/fortran/trans-openmp.c index d1799d7ebc21..5d0d6d2c011c 100644 --- a/gcc/fortran/trans-openmp.c +++ b/gcc/fortran/trans-openmp.c @@ -2280,7 +2280,7 @@ gfc_trans_omp_clauses (stmtblock_t *block, gfc_omp_clauses *clauses, tree node = build_omp_clause (input_location, OMP_CLAUSE_DEPEND); if (n->expr == NULL || n->expr->ref->u.ar.type == AR_FULL) { - tree decl = gfc_get_symbol_decl (n->sym); + tree decl = gfc_trans_omp_variable (n->sym, false); if (gfc_omp_privatize_by_reference (decl)) decl = build_fold_indirect_ref (decl); if (GFC_DESCRIPTOR_TYPE_P (TREE_TYPE (decl))) @@ -2341,7 +2341,7 @@ gfc_trans_omp_clauses (stmtblock_t *block, gfc_omp_clauses *clauses, tree node2 = NULL_TREE; tree node3 = NULL_TREE; tree node4 = NULL_TREE; - tree decl = gfc_get_symbol_decl (n->sym); + tree decl = gfc_trans_omp_variable (n->sym, false); if (DECL_P (decl)) TREE_ADDRESSABLE (decl) = 1; @@ -2749,7 +2749,7 @@ gfc_trans_omp_clauses (stmtblock_t *block, gfc_omp_clauses *clauses, tree node = build_omp_clause (input_location, clause_code); if (n->expr == NULL || n->expr->ref->u.ar.type == AR_FULL) { - tree decl = gfc_get_symbol_decl (n->sym); + tree decl = gfc_trans_omp_variable (n->sym, false); if (gfc_omp_privatize_by_reference (decl)) decl = build_fold_indirect_ref (decl); else if (DECL_P (decl)) diff --git a/gcc/testsuite/ChangeLog.omp b/gcc/testsuite/ChangeLog.omp index 5f452e64a93d..8282d962fc90 100644 --- a/gcc/testsuite/ChangeLog.omp +++ b/gcc/testsuite/ChangeLog.omp @@ -1,3 +1,18 @@ +2019-09-20 Tobias Burnus + + Backported from mainline + 2019-09-20 Tobias Burnus + + * gfortran.dg/goacc/parameter.f95: Change + dg-error as it is now detected earlier. + * gfortran.dg/goacc/pr85701.f90: Modify to + use a separate result variable. + * gfortran.dg/goacc/pr78260.f90: New. + * gfortran.dg/goacc/pr78260-2.f90: New. + * gfortran.dg/gomp/pr78260.f90: New. + * gfortran.dg/gomp/pr78260-2.f90: New. + * gfortran.dg/gomp/pr78260-3.f90: New. + 2019-09-19 Tobias Burnus * gfortran.dg/goacc/classify-kernels-unparallelized.f95: Add diff --git a/gcc/testsuite/gfortran.dg/goacc/parameter.f95 b/gcc/testsuite/gfortran.dg/goacc/parameter.f95 index 84274611915c..cbe67dba7889 100644 --- a/gcc/testsuite/gfortran.dg/goacc/parameter.f95 +++ b/gcc/testsuite/gfortran.dg/goacc/parameter.f95 @@ -6,7 +6,7 @@ contains implicit none integer :: i integer, parameter :: a = 1 - !$acc declare device_resident (a) ! { dg-error "PARAMETER" } + !$acc declare device_resident (a) ! { dg-error "is not a variable" } !$acc data copy (a) ! { dg-error "not a variable" } !$acc end data !$acc data deviceptr (a) ! { dg-error "not a variable" } diff --git a/gcc/testsuite/gfortran.dg/goacc/pr78260-2.f90 b/gcc/testsuite/gfortran.dg/goacc/pr78260-2.f90 new file mode 100644 index 000000000000..e28564d6f700 --- /dev/null +++ b/gcc/testsuite/gfortran.dg/goacc/pr78260-2.f90 @@ -0,0 +1,20 @@ +! { dg-do compile } +! { dg-options "-fopenacc -fdump-tree-original" } +! { dg-require-effective-target fopenacc } + +! PR fortran/78260 + +module m + implicit none + integer :: n = 0 +contains + integer function f1() + !$acc declare present(f1) + !$acc kernels copyin(f1) + f1 = 5 + !$acc end kernels + end function f1 +end module m +! { dg-final { scan-tree-dump-times "#pragma acc data map\\(force_present:__result_f1\\)" 1 "original" } } +! { dg-final { scan-tree-dump-times "#pragma acc data map\\(force_present:__result_f1\\)" 1 "original" } } + diff --git a/gcc/testsuite/gfortran.dg/goacc/pr78260.f90 b/gcc/testsuite/gfortran.dg/goacc/pr78260.f90 new file mode 100644 index 000000000000..21bde854919d --- /dev/null +++ b/gcc/testsuite/gfortran.dg/goacc/pr78260.f90 @@ -0,0 +1,36 @@ +! { dg-do compile } +! { dg-options "-fopenacc" } +! { dg-require-effective-target fopenacc } + +! PR fortran/78260 +! Contributed by Gerhard Steinmetz + +module m + implicit none + integer :: n = 0 +contains + subroutine s + !$acc declare present(m) ! { dg-error "Object .m. is not a variable" } + !$acc kernels copyin(m) ! { dg-error "Object .m. is not a variable" } + n = n + 1 + !$acc end kernels + end subroutine s + subroutine s2 + !$acc declare present(s2) ! { dg-error "Object .s2. is not a variable" } + !$acc kernels copyin(s2) ! { dg-error "Object .s2. is not a variable" } + n = n + 1 + !$acc end kernels + end subroutine s2 + integer function f1() + !$acc declare present(f1) ! OK, f1 is also the result variable + !$acc kernels copyin(f1) ! OK, f1 is also the result variable + f1 = 5 + !$acc end kernels + end function f1 + integer function f2() result(res) + !$acc declare present(f2) ! { dg-error "Object .f2. is not a variable" } + !$acc kernels copyin(f2) ! { dg-error "Object .f2. is not a variable" } + res = 5 + !$acc end kernels + end function f2 +end module m diff --git a/gcc/testsuite/gfortran.dg/goacc/pr85701.f90 b/gcc/testsuite/gfortran.dg/goacc/pr85701.f90 index 9c201b865b28..bae09de90aca 100644 --- a/gcc/testsuite/gfortran.dg/goacc/pr85701.f90 +++ b/gcc/testsuite/gfortran.dg/goacc/pr85701.f90 @@ -9,11 +9,11 @@ subroutine s2 !$acc declare present(s2) ! { dg-error "is not a variable" } end -function f1 () +function f1 () result(res) !$acc declare copy(f1) ! { dg-error "is not a variable" } end -function f2 () +function f2 () result(res) !$acc declare present(f2) ! { dg-error "is not a variable" } end diff --git a/gcc/testsuite/gfortran.dg/gomp/pr78260-2.f90 b/gcc/testsuite/gfortran.dg/gomp/pr78260-2.f90 new file mode 100644 index 000000000000..c58ad93471c2 --- /dev/null +++ b/gcc/testsuite/gfortran.dg/gomp/pr78260-2.f90 @@ -0,0 +1,59 @@ +! { dg-do compile } +! { dg-options "-fopenmp -fdump-tree-original" } + +! PR fortran/78260 + +module m + implicit none + integer :: n = 0 +contains + integer function f1() + !$omp target data map(f1) + !$omp target update to(f1) + f1 = 5 + !$omp end target data + end function f1 + + integer function f2() + dimension :: f2(1) + !$omp target data map(f2) + !$omp target update to(f2) + f2(1) = 5 + !$omp end target data + end function f2 + + integer function f3() result(res) + dimension :: res(1) + !$omp target data map(res) + !$omp target update to(res) + res(1) = 5 + !$omp end target data + end function f3 + + integer function f4() result(res) + allocatable :: res + dimension :: res(:) + !$omp target data map(res) + !$omp target update to(res) + res = [5] + !$omp end target data + end function f4 + + subroutine sub() + integer, allocatable :: arr(:) + !$omp target data map(arr) + !$omp target update to(arr) + arr = [5] + !$omp end target data + end subroutine sub +end module m + +! { dg-final { scan-tree-dump-times "#pragma omp target data map\\(tofrom:\\*\\(c_char \\*\\) arr.data \\\[len: D.\[0-9\]+ \\* 4\\\]\\) map\\(to:arr \\\[pointer set, len: ..\\\]\\) map\\(alloc:\\(integer\\(kind=4\\)\\\[0:\\\] \\* restrict\\) arr.data \\\[pointer assign, bias: 0\\\]\\)" 1 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp target update to\\(\\*\\(c_char \\*\\) arr.data \\\[len: D.\[0-9\]+ \\* 4\\\]\\)" 1 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp target data map\\(tofrom:\\*\\(c_char \\*\\) __result->data \\\[len: D.\[0-9\]+ \\* 4\\\]\\) map\\(to:\\*__result \\\[pointer set, len: ..\\\]\\) map\\(alloc:\\(integer\\(kind=4\\)\\\[0:\\\] \\* restrict\\) __result->data \\\[pointer assign, bias: 0\\\]\\) map\\(alloc:__result \\\[pointer assign, bias: 0\\\]\\)" 1 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp target update to\\(\\*\\(c_char \\*\\) __result->data \\\[len: D.\[0-9\]+ \\* 4\\\]\\)" 1 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp target data map\\(tofrom:\\*__result.0\\) map\\(alloc:__result.0 \\\[pointer assign, bias: 0\\\]\\)" 2 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp target update to\\(\\*__result.0\\)" 2 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp target data map\\(tofrom:__result_f1\\)" 1 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp target update to\\(__result_f1\\)" 1 "original" } } + diff --git a/gcc/testsuite/gfortran.dg/gomp/pr78260-3.f90 b/gcc/testsuite/gfortran.dg/gomp/pr78260-3.f90 new file mode 100644 index 000000000000..4ca3e361a59b --- /dev/null +++ b/gcc/testsuite/gfortran.dg/gomp/pr78260-3.f90 @@ -0,0 +1,74 @@ +! { dg-do compile } +! { dg-options "-fopenmp -fdump-tree-original" } + +! PR fortran/78260 + +integer function f1() + implicit none + + f1 = 0 + + !$omp task depend(inout:f1) + !$omp end task + + !$omp task depend(inout:f1) + !$omp end task +end function f1 + +integer function f2() + implicit none + dimension :: f2(1) + + f2(1) = 0 + + !$omp task depend(inout:f2) + !$omp end task + + !$omp task depend(inout:f2) + !$omp end task +end function f2 + +integer function f3() result(res) + implicit none + dimension :: res(1) + + res(1) = 0 + + !$omp task depend(inout:res) + !$omp end task + + !$omp task depend(inout:res) + !$omp end task +end function f3 + +integer function f4() result(res) + implicit none + allocatable :: res + dimension :: res(:) + + res = [0] + + !$omp task depend(inout:res) + !$omp end task + + !$omp task depend(inout:res) + !$omp end task +end function f4 + +subroutine sub() + implicit none + integer, allocatable :: arr(:) + + arr = [3] + + !$omp task depend(inout:arr) + !$omp end task + + !$omp task depend(inout:arr) + !$omp end task +end subroutine sub + +! { dg-final { scan-tree-dump-times "#pragma omp task depend\\(inout:__result_f1\\)" 2 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp task depend\\(inout:\\*__result.0\\)" 4 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp task depend\\(inout:\\*\\(c_char \\*\\) __result->data\\)" 2 "original" } } +! { dg-final { scan-tree-dump-times "#pragma omp task depend\\(inout:\\*\\(c_char \\*\\) arr.data\\)" 2 "original" } } diff --git a/gcc/testsuite/gfortran.dg/gomp/pr78260.f90 b/gcc/testsuite/gfortran.dg/gomp/pr78260.f90 new file mode 100644 index 000000000000..23acd4c1bf92 --- /dev/null +++ b/gcc/testsuite/gfortran.dg/gomp/pr78260.f90 @@ -0,0 +1,33 @@ +! { dg-do compile } + +! PR fortran/78260 + +module m + implicit none + integer :: n = 0 +contains + subroutine s + !$omp target data map(m) ! { dg-error "Object .m. is not a variable" } + !$omp target update to(m) ! { dg-error "Object .m. is not a variable" } + n = n + 1 + !$omp end target data + end subroutine s + subroutine s2 + !$omp target data map(s2) ! { dg-error "Object .s2. is not a variable" } + !$omp target update to(s2) ! { dg-error "Object .s2. is not a variable" } + n = n + 1 + !$omp end target data + end subroutine s2 + integer function f1() + !$omp target data map(f1) ! OK, f1 is also the result variable + !$omp target update to(f1) ! OK, f1 is also the result variable + f1 = 5 + !$omp end target data + end function f1 + integer function f2() result(res) + !$omp target data map(f2) ! { dg-error "Object .f2. is not a variable" } + !$omp target update to(f2) ! { dg-error "Object .f2. is not a variable" } + res = 5 + !$omp end target data + end function f2 +end module m