public inbox for gcc-patches@gcc.gnu.org
 help / color / mirror / Atom feed
* [PATCH, OpenACC 2.7] Implement host_data must have use_device clause requirement
@ 2023-06-06 15:10 Chung-Lin Tang
  2023-06-16  9:13 ` Thomas Schwinge
  0 siblings, 1 reply; 4+ messages in thread
From: Chung-Lin Tang @ 2023-06-06 15:10 UTC (permalink / raw)
  To: gcc-patches, Thomas Schwinge, Catherine Moore

[-- Attachment #1: Type: text/plain, Size: 1238 bytes --]

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.

[-- Attachment #2: 0001-OpenACC-2.7-host_data-must-have-use_device-clause-re.patch --]
[-- Type: text/plain, Size: 5694 bytes --]

From 0d17b8d24fa6079d6c289305e9644c3fecd429f1 Mon Sep 17 00:00:00 2001
From: Chung-Lin Tang <cltang@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


^ permalink raw reply	[flat|nested] 4+ messages in thread

* Re: [PATCH, OpenACC 2.7] Implement host_data must have use_device clause requirement
  2023-06-06 15:10 [PATCH, OpenACC 2.7] Implement host_data must have use_device clause requirement Chung-Lin Tang
@ 2023-06-16  9:13 ` Thomas Schwinge
  2023-07-13 10:54   ` [PATCH, OpenACC 2.7, v2] " Chung-Lin Tang
  0 siblings, 1 reply; 4+ messages in thread
From: Thomas Schwinge @ 2023-06-16  9:13 UTC (permalink / raw)
  To: Chung-Lin Tang; +Cc: gcc-patches, Catherine Moore

Hi Chung-Lin!

On 2023-06-06T23:10:37+0800, Chung-Lin Tang <chunglin.tang@siemens.com> wrote:
> this patch implements the OpenACC 2.7 change requiring the host_data construct
> to have at least one use_device clause.

Thanks!

> This patch started out with a simple check during gimplify (much smaller patch),

Heh, thanks for the explanation -- would've been my first question
otherweise.  ;-)

> 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.

I presume that's also the reason why you're doing this check before
'c_finish_omp_clauses' etc.?

I'll clarify with the OpenACC Technical Committee whether really those
diagnostics are intended as "error" or should instead just be "warning".
After all, there's no actual problem with an OpenACC 'host_data' without
'use_device' clause (or no data clause on OpenACC 'data', 'enter data',
'exit data', 'update', etc.) -- it's just likely that the user missed
something.  That is, the OpenACC 2.7: "At least one 'use_device' clause
must appear" is addressing the user, not at the implementation (in my
current interpretation).  Depending on the outcome of that, we can easily
adjust GCC.

Note for later, independently of your work here:
'c_parser_oacc_enter_exit_data' etc. for its corresponding "has no data
movement clause" diagnostic actually does 'c_finish_omp_clauses' etc.
first -- maybe that should be changed accordingly.  (Actually, I note
that it's only OpenACC 3.0 that "Required at least one data clause on a
'data' construct, an 'enter data' directive, or an 'exit data'
directive", heh...  Per his internal 2014-10-17 email, Cesar implemented
the code of 'c_parser_oacc_enter_exit_data' etc. "similar to that of acc
update", which indeed already back then did require "At least one 'self',
'host', or 'device' clause".  Fortran does have the diagnostic for
OpenACC 'update', but it's missing for OpenACC 'enter data', 'exit data'
without data clause (have not checked other constructs with similar
requirements).)

> Tested on powerpc64le-linux/nvptx, x86_64-linux/amdgcn tests in progress (expect
> no surprises). Is this okay for trunk?

OK with one small change, please -- unless there's a reason for doing it
this way:

> --- 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 ();

The OpenMP "must contain at least one [...] clause" checks are done in
'gcc/fortran/openmp.cc:resolve_omp_clauses'.  For consistency (or, to let
'gcc/fortran/trans-openmp.cc' continue to just deal with "directive
translation"), do similar for OpenACC 'host_data'?  (..., and we later
accordingly adjust 'gcc/fortran/openmp.cc:gfc_match_oacc_update', too?)


Grüße
 Thomas
-----------------
Siemens Electronic Design Automation GmbH; Anschrift: Arnulfstraße 201, 80634 München; Gesellschaft mit beschränkter Haftung; Geschäftsführer: Thomas Heurung, Frank Thürauf; Sitz der Gesellschaft: München; Registergericht München, HRB 106955

^ permalink raw reply	[flat|nested] 4+ messages in thread

* [PATCH, OpenACC 2.7, v2] Implement host_data must have use_device clause requirement
  2023-06-16  9:13 ` Thomas Schwinge
@ 2023-07-13 10:54   ` Chung-Lin Tang
  2023-07-20 10:00     ` Thomas Schwinge
  0 siblings, 1 reply; 4+ messages in thread
From: Chung-Lin Tang @ 2023-07-13 10:54 UTC (permalink / raw)
  To: Thomas Schwinge, Chung-Lin Tang; +Cc: gcc-patches, Catherine Moore

[-- Attachment #1: Type: text/plain, Size: 1927 bytes --]

On 2023/6/16 5:13 PM, Thomas Schwinge wrote:
> OK with one small change, please -- unless there's a reason for doing it
> this way:
> 
>> --- 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 ();
> The OpenMP "must contain at least one [...] clause" checks are done in
> 'gcc/fortran/openmp.cc:resolve_omp_clauses'.  For consistency (or, to let
> 'gcc/fortran/trans-openmp.cc' continue to just deal with "directive
> translation"), do similar for OpenACC 'host_data'?  (..., and we later
> accordingly adjust 'gcc/fortran/openmp.cc:gfc_match_oacc_update', too?)

Hi Thomas,
I've adjusted the Fortran implementation as you described. Yes, I agree this way
more fits current Fortran FE conventions.

I've re-tested the attached v2 patch, will commit later this week if no major
objections.

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:

	* openmp.cc (resolve_omp_clauses): 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.

[-- Attachment #2: host_data-use_device-req-v2.patch --]
[-- Type: text/plain, Size: 4212 bytes --]

diff --git a/gcc/c/c-parser.cc b/gcc/c/c-parser.cc
index 24a6eb6e459..80920b31f83 100644
--- a/gcc/c/c-parser.cc
+++ b/gcc/c/c-parser.cc
@@ -18461,8 +18461,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 5e2b5cba57e..beb5b632e5e 100644
--- a/gcc/cp/parser.cc
+++ b/gcc/cp/parser.cc
@@ -45895,8 +45895,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/openmp.cc b/gcc/fortran/openmp.cc
index 8efc4b3ecfa..f7af02845de 100644
--- a/gcc/fortran/openmp.cc
+++ b/gcc/fortran/openmp.cc
@@ -8764,6 +8764,12 @@ resolve_omp_clauses (gfc_code *code, gfc_omp_clauses *omp_clauses,
 		   "%<MERGEABLE%> clause", &omp_clauses->detach->where);
     }
 
+  if (openacc
+      && code->op == EXEC_OACC_HOST_DATA
+      && omp_clauses->lists[OMP_LIST_USE_DEVICE] == NULL)
+    gfc_error ("%<host_data%> construct at %L requires %<use_device%> clause",
+	       &code->loc);
+
   if (omp_clauses->assume)
     gfc_resolve_omp_assumptions (omp_clauses->assume);
 }
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..bd262989410
--- /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 at .1. 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

^ permalink raw reply	[flat|nested] 4+ messages in thread

* Re: [PATCH, OpenACC 2.7, v2] Implement host_data must have use_device clause requirement
  2023-07-13 10:54   ` [PATCH, OpenACC 2.7, v2] " Chung-Lin Tang
@ 2023-07-20 10:00     ` Thomas Schwinge
  0 siblings, 0 replies; 4+ messages in thread
From: Thomas Schwinge @ 2023-07-20 10:00 UTC (permalink / raw)
  To: Chung-Lin Tang; +Cc: gcc-patches, Catherine Moore

Hi Chung-Lin!

On 2023-07-13T18:54:00+0800, Chung-Lin Tang <chunglin.tang@siemens.com> wrote:
> On 2023/6/16 5:13 PM, Thomas Schwinge wrote:
>> OK with one small change, please -- unless there's a reason for doing it
>> this way: [...]

> I've adjusted the Fortran implementation as you described. Yes, I agree this way
> more fits current Fortran FE conventions.
>
> I've re-tested the attached v2 patch, will commit later this week if no major
> objections.

ACK, thanks.


Grüße
 Thomas


> 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:
>
>       * openmp.cc (resolve_omp_clauses): 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.
> diff --git a/gcc/c/c-parser.cc b/gcc/c/c-parser.cc
> index 24a6eb6e459..80920b31f83 100644
> --- a/gcc/c/c-parser.cc
> +++ b/gcc/c/c-parser.cc
> @@ -18461,8 +18461,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 5e2b5cba57e..beb5b632e5e 100644
> --- a/gcc/cp/parser.cc
> +++ b/gcc/cp/parser.cc
> @@ -45895,8 +45895,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/openmp.cc b/gcc/fortran/openmp.cc
> index 8efc4b3ecfa..f7af02845de 100644
> --- a/gcc/fortran/openmp.cc
> +++ b/gcc/fortran/openmp.cc
> @@ -8764,6 +8764,12 @@ resolve_omp_clauses (gfc_code *code, gfc_omp_clauses *omp_clauses,
>                  "%<MERGEABLE%> clause", &omp_clauses->detach->where);
>      }
>
> +  if (openacc
> +      && code->op == EXEC_OACC_HOST_DATA
> +      && omp_clauses->lists[OMP_LIST_USE_DEVICE] == NULL)
> +    gfc_error ("%<host_data%> construct at %L requires %<use_device%> clause",
> +            &code->loc);
> +
>    if (omp_clauses->assume)
>      gfc_resolve_omp_assumptions (omp_clauses->assume);
>  }
> 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..bd262989410
> --- /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 at .1. 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
-----------------
Siemens Electronic Design Automation GmbH; Anschrift: Arnulfstraße 201, 80634 München; Gesellschaft mit beschränkter Haftung; Geschäftsführer: Thomas Heurung, Frank Thürauf; Sitz der Gesellschaft: München; Registergericht München, HRB 106955

^ permalink raw reply	[flat|nested] 4+ messages in thread

end of thread, other threads:[~2023-07-20 10:00 UTC | newest]

Thread overview: 4+ messages (download: mbox.gz / follow: Atom feed)
-- links below jump to the message on this page --
2023-06-06 15:10 [PATCH, OpenACC 2.7] Implement host_data must have use_device clause requirement Chung-Lin Tang
2023-06-16  9:13 ` Thomas Schwinge
2023-07-13 10:54   ` [PATCH, OpenACC 2.7, v2] " Chung-Lin Tang
2023-07-20 10:00     ` Thomas Schwinge

This is a public inbox, see mirroring instructions
for how to clone and mirror all data and code used for this inbox;
as well as URLs for read-only IMAP folder(s) and NNTP newsgroup(s).