Hi Thomas,
this patch implements the OpenACC 2.7 change requiring the host_data construct
to have at least one use_device clause.
This patch started out with a simple check during gimplify (much smaller patch),
but turned out that front-ends removed use_device clauses when they have error,
and the gimplify check started to echo a "no use_device clause" message in such
cases, which seem confusing for the user. So ended up adding the check in each
front-end instead.
Tested on powerpc64le-linux/nvptx, x86_64-linux/amdgcn tests in progress (expect
no surprises). Is this okay for trunk?
Thanks,
Chung-Lin
gcc/c/ChangeLog:
* c-parser.cc (c_parser_oacc_host_data): Add checking requiring OpenACC
host_data construct to have an use_device clause.
gcc/cp/ChangeLog:
* parser.cc (cp_parser_oacc_host_data): Add checking requiring OpenACC
host_data construct to have an use_device clause.
gcc/fortran/ChangeLog:
* trans-openmp.cc (gfc_trans_oacc_construct): Add checking requiring
OpenACC host_data construct to have an use_device clause.
gcc/testsuite/ChangeLog:
* c-c++-common/goacc/host_data-2.c: Adjust testcase.
* gfortran.dg/goacc/host_data-error.f90: New testcase.
* gfortran.dg/goacc/pr71704.f90: Adjust testcase.
From 0d17b8d24fa6079d6c289305e9644c3fecd429f1 Mon Sep 17 00:00:00 2001
From: Chung-Lin Tang <clt...@codesourcery.com>
Date: Tue, 6 Jun 2023 03:19:33 -0700
Subject: [PATCH 1/2] OpenACC 2.7: host_data must have use_device clause
requirement
This patch implements the OpenACC 2.7 change requiring the host_data construct
to have at least one use_device clause.
gcc/c/ChangeLog:
* c-parser.cc (c_parser_oacc_host_data): Add checking requiring OpenACC
host_data construct to have an use_device clause.
gcc/cp/ChangeLog:
* parser.cc (cp_parser_oacc_host_data): Add checking requiring OpenACC
host_data construct to have an use_device clause.
gcc/fortran/ChangeLog:
* trans-openmp.cc (gfc_trans_oacc_construct): Add checking requiring
OpenACC host_data construct to have an use_device clause.
gcc/testsuite/ChangeLog:
* c-c++-common/goacc/host_data-2.c: Adjust testcase.
* gfortran.dg/goacc/host_data-error.f90: New testcase.
* gfortran.dg/goacc/pr71704.f90: Adjust testcase.
---
gcc/c/c-parser.cc | 9 +++++++--
gcc/cp/parser.cc | 11 +++++++++--
gcc/fortran/trans-openmp.cc | 6 ++++++
gcc/testsuite/c-c++-common/goacc/host_data-2.c | 7 ++++++-
gcc/testsuite/gfortran.dg/goacc/host_data-error.f90 | 6 ++++++
gcc/testsuite/gfortran.dg/goacc/pr71704.f90 | 5 +++--
6 files changed, 37 insertions(+), 7 deletions(-)
create mode 100644 gcc/testsuite/gfortran.dg/goacc/host_data-error.f90
diff --git a/gcc/c/c-parser.cc b/gcc/c/c-parser.cc
index 5baa501dbee..b61aef8b1a2 100644
--- a/gcc/c/c-parser.cc
+++ b/gcc/c/c-parser.cc
@@ -18398,8 +18398,13 @@ c_parser_oacc_host_data (location_t loc, c_parser
*parser, bool *if_p)
tree stmt, clauses, block;
clauses = c_parser_oacc_all_clauses (parser, OACC_HOST_DATA_CLAUSE_MASK,
- "#pragma acc host_data");
-
+ "#pragma acc host_data", false);
+ if (!omp_find_clause (clauses, OMP_CLAUSE_USE_DEVICE_PTR))
+ {
+ error_at (loc, "%<host_data%> construct requires %<use_device%> clause");
+ return error_mark_node;
+ }
+ clauses = c_finish_omp_clauses (clauses, C_ORT_ACC);
block = c_begin_omp_parallel ();
add_stmt (c_parser_omp_structured_block (parser, if_p));
stmt = c_finish_oacc_host_data (loc, clauses, block);
diff --git a/gcc/cp/parser.cc b/gcc/cp/parser.cc
index 1c9aa671851..dd7638f1c93 100644
--- a/gcc/cp/parser.cc
+++ b/gcc/cp/parser.cc
@@ -45798,8 +45798,15 @@ cp_parser_oacc_host_data (cp_parser *parser, cp_token
*pragma_tok, bool *if_p)
unsigned int save;
clauses = cp_parser_oacc_all_clauses (parser, OACC_HOST_DATA_CLAUSE_MASK,
- "#pragma acc host_data", pragma_tok);
-
+ "#pragma acc host_data", pragma_tok,
+ false);
+ if (!omp_find_clause (clauses, OMP_CLAUSE_USE_DEVICE_PTR))
+ {
+ error_at (pragma_tok->location,
+ "%<host_data%> construct requires %<use_device%> clause");
+ return error_mark_node;
+ }
+ clauses = finish_omp_clauses (clauses, C_ORT_ACC);
block = begin_omp_parallel ();
save = cp_parser_begin_omp_structured_block (parser);
cp_parser_statement (parser, NULL_TREE, false, if_p);
diff --git a/gcc/fortran/trans-openmp.cc b/gcc/fortran/trans-openmp.cc
index 42b608f3d36..5e0079cce76 100644
--- a/gcc/fortran/trans-openmp.cc
+++ b/gcc/fortran/trans-openmp.cc
@@ -4677,6 +4677,12 @@ gfc_trans_oacc_construct (gfc_code *code)
break;
case EXEC_OACC_HOST_DATA:
construct_code = OACC_HOST_DATA;
+ if (code->ext.omp_clauses->lists[OMP_LIST_USE_DEVICE] == NULL)
+ {
+ error_at (gfc_get_location (&code->loc),
+ "%<host_data%> construct requires %<use_device%> clause");
+ return NULL_TREE;
+ }
break;
default:
gcc_unreachable ();
diff --git a/gcc/testsuite/c-c++-common/goacc/host_data-2.c
b/gcc/testsuite/c-c++-common/goacc/host_data-2.c
index b3093e575ff..862a764eb3a 100644
--- a/gcc/testsuite/c-c++-common/goacc/host_data-2.c
+++ b/gcc/testsuite/c-c++-common/goacc/host_data-2.c
@@ -8,7 +8,9 @@ void
f (void)
{
int v2 = 3;
-#pragma acc host_data copy(v2) /* { dg-error ".copy. is not valid for ..pragma
acc host_data." } */
+#pragma acc host_data copy(v2)
+ /* { dg-error ".copy. is not valid for ..pragma acc host_data." "" { target
*-*-* } .-1 } */
+ /* { dg-error ".host_data. construct requires .use_device. clause" "" {
target *-*-* } .-2 } */
;
#pragma acc host_data use_device(v2)
@@ -20,6 +22,9 @@ f (void)
/* { dg-error ".use_device_ptr. variable is neither a pointer nor an array"
"" { target c } .-1 } */
/* { dg-error ".use_device_ptr. variable is neither a pointer, nor an array
nor reference to pointer or array" "" { target c++ } .-2 } */
;
+
+#pragma acc host_data /* { dg-error ".host_data. construct requires
.use_device. clause" } */
+ ;
}
diff --git a/gcc/testsuite/gfortran.dg/goacc/host_data-error.f90
b/gcc/testsuite/gfortran.dg/goacc/host_data-error.f90
new file mode 100644
index 00000000000..747a2b201e7
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/goacc/host_data-error.f90
@@ -0,0 +1,6 @@
+! { dg-do compile }
+
+subroutine foo ()
+!$acc host_data ! { dg-error "'host_data' construct requires 'use_device'
clause" }
+!$acc end host_data
+end
diff --git a/gcc/testsuite/gfortran.dg/goacc/pr71704.f90
b/gcc/testsuite/gfortran.dg/goacc/pr71704.f90
index 0235e85d42a..31724c8b046 100644
--- a/gcc/testsuite/gfortran.dg/goacc/pr71704.f90
+++ b/gcc/testsuite/gfortran.dg/goacc/pr71704.f90
@@ -47,8 +47,9 @@ real function f8 ()
f8 = 1
end
-real function f9 ()
-!$acc host_data
+real function f9 (a)
+ integer a(:)
+!$acc host_data use_device(a)
!$acc end host_data
f8 = 1
end
--
2.27.0
