From c319667adff8cc10e87ff836f72d5a7471e942c6 Mon Sep 17 00:00:00 2001 From: Thomas Schwinge Date: Fri, 22 Feb 2019 11:50:18 +0100 Subject: [PATCH] Revert earlier OpenACC 'routine' changes ..., which were committed as part of, but logically don't belong into r261813 "Update OpenACC data clause semantics to the 2.5 behavior", and which confuse later changes. gcc/fortran/ * openmp.c (gfc_match_oacc_declare): Revert earlier changes. gcc/testsuite/ * c-c++-common/goacc/routine-5.c: Revert earlier changes. * g++.dg/goacc/template.C: Likewise. From-SVN: r269104 --- gcc/fortran/ChangeLog | 4 + gcc/fortran/openmp.c | 7 - gcc/testsuite/ChangeLog | 3 + gcc/testsuite/c-c++-common/goacc/routine-5.c | 150 ++++++++++++------- gcc/testsuite/g++.dg/goacc/template.C | 2 +- 5 files changed, 104 insertions(+), 62 deletions(-) diff --git a/gcc/fortran/ChangeLog b/gcc/fortran/ChangeLog index c3bc30c42f6..74a6890ed70 100644 --- a/gcc/fortran/ChangeLog +++ b/gcc/fortran/ChangeLog @@ -1,3 +1,7 @@ +2019-02-22 Thomas Schwinge + + * openmp.c (gfc_match_oacc_declare): Revert earlier changes. + 2019-02-21 Thomas Koenig * dump-parse-tree.c (debug): Implement for gfc_expr *, diff --git a/gcc/fortran/openmp.c b/gcc/fortran/openmp.c index 11dd3ae0ada..8aa4a2f18c4 100644 --- a/gcc/fortran/openmp.c +++ b/gcc/fortran/openmp.c @@ -2075,13 +2075,6 @@ gfc_match_oacc_declare (void) module_var = true; } - if (ns->proc_name->attr.oacc_function) - { - gfc_error ("Invalid declare in routine with $!ACC DECLARE at %L", - &where); - return MATCH_ERROR; - } - if (s->attr.use_assoc) { gfc_error ("Variable is USE-associated with !$ACC DECLARE at %L", diff --git a/gcc/testsuite/ChangeLog b/gcc/testsuite/ChangeLog index b98cd355645..12901a9361a 100644 --- a/gcc/testsuite/ChangeLog +++ b/gcc/testsuite/ChangeLog @@ -1,5 +1,8 @@ 2019-02-22 Thomas Schwinge + * c-c++-common/goacc/routine-5.c: Revert earlier changes. + * g++.dg/goacc/template.C: Likewise. + PR fortran/78027 * gfortran.dg/goacc/pr78027.f90: Add 'dg-additional-options "-Wno-hsa"'. diff --git a/gcc/testsuite/c-c++-common/goacc/routine-5.c b/gcc/testsuite/c-c++-common/goacc/routine-5.c index b759db3292d..b967a7447bd 100644 --- a/gcc/testsuite/c-c++-common/goacc/routine-5.c +++ b/gcc/testsuite/c-c++-common/goacc/routine-5.c @@ -4,11 +4,11 @@ struct PC { -#pragma acc routine seq /* { dg-error ".#pragma acc routine. must be at file scope" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */ }; void PC1( /* { dg-bogus "variable or field .PC1. declared void" "TODO" { xfail c++ } } */ -#pragma acc routine seq +#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" "" { target c } .-1 } { dg-error ".#pragma. is not allowed here" "" { target c++ } .-2 } */ ) /* { dg-bogus "expected declaration specifiers or .\\.\\.\\.. before .\\). token" "TODO" { xfail c } } */ @@ -18,26 +18,26 @@ void PC1( /* { dg-bogus "variable or field .PC1. declared void" "TODO" { xfail c void PC2() { if (0) -#pragma acc routine seq /* { dg-error ".#pragma acc routine. must be at file scope" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */ ; } void PC3() { -#pragma acc routine seq /* { dg-error ".#pragma acc routine. must be at file scope" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */ } /* "( name )" syntax. */ #pragma acc routine ( /* { dg-error "expected (function name|unqualified-id) before end of line" } */ -#pragma acc routine () seq /* { dg-error "expected (function name|unqualified-id) before .\\). token" } */ -#pragma acc routine (+) seq /* { dg-error "expected (function name|unqualified-id) before .\\+. token" } */ -#pragma acc routine (?) seq /* { dg-error "expected (function name|unqualified-id) before .\\?. token" } */ -#pragma acc routine (:) seq /* { dg-error "expected (function name|unqualified-id) before .:. token" } */ -#pragma acc routine (4) seq /* { dg-error "expected (function name|unqualified-id) before numeric constant" } */ +#pragma acc routine () /* { dg-error "expected (function name|unqualified-id) before .\\). token" } */ +#pragma acc routine (+) /* { dg-error "expected (function name|unqualified-id) before .\\+. token" } */ +#pragma acc routine (?) /* { dg-error "expected (function name|unqualified-id) before .\\?. token" } */ +#pragma acc routine (:) /* { dg-error "expected (function name|unqualified-id) before .:. token" } */ +#pragma acc routine (4) /* { dg-error "expected (function name|unqualified-id) before numeric constant" } */ #pragma acc routine ('4') /* { dg-error "expected (function name|unqualified-id) before .4." } */ -#pragma acc routine ("4") seq /* { dg-error "expected (function name|unqualified-id) before string constant" } */ +#pragma acc routine ("4") /* { dg-error "expected (function name|unqualified-id) before string constant" } */ extern void R1(void); extern void R2(void); #pragma acc routine (R1, R2, R3) worker /* { dg-error "expected .\\). before .,. token" } */ @@ -49,84 +49,84 @@ extern void R2(void); /* "#pragma acc routine" not immediately followed by (a single) function declaration or definition. */ -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ int a; -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ void fn1 (void), fn1b (void); -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ int b, fn2 (void); -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ int b_, fn2_ (void), B_; -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ int fn3 (void), b2; -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ typedef struct c c; -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ struct d {} d; -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ void fn1_2 (void), fn1b_2 (void); -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ #pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ int b_2, fn2_2 (void); -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ int b_2_, fn2_2_ (void), B_2_; -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ int fn3_2 (void), b2_2; -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ typedef struct c_2 c_2; -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ struct d_2 {} d_2; -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ -#pragma acc routine seq +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine int fn4 (void); int fn5a (void); int fn5b (void); -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ -#pragma acc routine (fn5a) seq -#pragma acc routine (fn5b) seq +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine (fn5a) +#pragma acc routine (fn5b) int fn5 (void); -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ -#pragma acc routine (fn6a) seq /* { dg-error ".fn6a. has not been declared" } */ -#pragma acc routine (fn6b) seq /* { dg-error ".fn6b. has not been declared" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine (fn6a) /* { dg-error ".fn6a. has not been declared" } */ +#pragma acc routine (fn6b) /* { dg-error ".fn6b. has not been declared" } */ int fn6 (void); #ifdef __cplusplus -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */ namespace f {} namespace g {} -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */ using namespace g; -#pragma acc routine (g) seq /* { dg-error ".g. does not refer to a function" "" { target c++ } } */ +#pragma acc routine (g) /* { dg-error ".g. does not refer to a function" "" { target c++ } } */ #endif /* __cplusplus */ -#pragma acc routine (a) seq /* { dg-error ".a. does not refer to a function" } */ +#pragma acc routine (a) /* { dg-error ".a. does not refer to a function" } */ -#pragma acc routine (c) seq /* { dg-error ".c. does not refer to a function" } */ +#pragma acc routine (c) /* { dg-error ".c. does not refer to a function" } */ /* Static assert. */ @@ -143,24 +143,66 @@ static_assert(0, ""); /* { dg-error "static assertion failed" "" { target c++11 #endif void f_static_assert(); /* Check that we already recognized "f_static_assert" as an OpenACC routine. */ -#pragma acc routine (f_static_assert) seq /* { dg-error ".#pragma acc routine. already applied to .\[void \]*f_static_assert" "TODO" { xfail *-*-* } } */ +#pragma acc routine (f_static_assert) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*f_static_assert" "TODO" { xfail *-*-* } } */ /* __extension__ usage. */ -#pragma acc routine seq +#pragma acc routine __extension__ extern void ex1(); #pragma acc routine (ex1) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*ex1" } */ -#pragma acc routine seq +#pragma acc routine __extension__ __extension__ __extension__ __extension__ __extension__ void ex2() { } #pragma acc routine (ex2) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*ex2" } */ -#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ __extension__ int ex3; -#pragma acc routine (ex3) seq /* { dg-error ".ex3. does not refer to a function" } */ +#pragma acc routine (ex3) /* { dg-error ".ex3. does not refer to a function" } */ + + +/* "#pragma acc routine" already applied. */ + +extern void fungsi_1(); +#pragma acc routine(fungsi_1) gang +#pragma acc routine(fungsi_1) gang /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_1" } */ +#pragma acc routine(fungsi_1) worker /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_1" } */ +#pragma acc routine(fungsi_1) vector /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_1" } */ + +#pragma acc routine seq +extern void fungsi_2(); +#pragma acc routine(fungsi_2) seq /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_2." } */ +#pragma acc routine(fungsi_2) worker /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_2." } */ +#pragma acc routine(fungsi_2) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_2." } */ + +#pragma acc routine vector +extern void fungsi_3(); +#pragma acc routine vector /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_3." } */ +void fungsi_3() +{ +} + +extern void fungsi_4(); +#pragma acc routine (fungsi_4) worker +#pragma acc routine gang /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_4." } */ +void fungsi_4() +{ +} + +#pragma acc routine gang +void fungsi_5() +{ +} +#pragma acc routine (fungsi_5) worker /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_5." } */ + +#pragma acc routine seq +void fungsi_6() +{ +} +#pragma acc routine seq /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_6." } */ +extern void fungsi_6(); /* "#pragma acc routine" must be applied before. */ @@ -172,11 +214,11 @@ void Foo () Bar (); } -#pragma acc routine (Bar) seq // { dg-error ".#pragma acc routine. must be applied before use" } +#pragma acc routine (Bar) // { dg-error ".#pragma acc routine. must be applied before use" } #pragma acc routine (Foo) gang // { dg-error ".#pragma acc routine. must be applied before definition" } -#pragma acc routine (Baz) seq // { dg-error "not been declared" } +#pragma acc routine (Baz) // { dg-error "not been declared" } /* OpenACC declare. */ @@ -185,7 +227,7 @@ int vb1; /* { dg-error "directive for use" } */ extern int vb2; /* { dg-error "directive for use" } */ static int vb3; /* { dg-error "directive for use" } */ -#pragma acc routine seq +#pragma acc routine int func1 (int a) { @@ -196,7 +238,7 @@ func1 (int a) return vb3; } -#pragma acc routine seq +#pragma acc routine int func2 (int a) { @@ -214,7 +256,7 @@ extern int vb6; /* { dg-error "clause used in" } */ static int vb7; /* { dg-error "clause used in" } */ #pragma acc declare link (vb7) -#pragma acc routine seq +#pragma acc routine int func3 (int a) { @@ -231,7 +273,7 @@ extern int vb9; static int vb10; #pragma acc declare create (vb10) -#pragma acc routine seq +#pragma acc routine int func4 (int a) { @@ -249,7 +291,7 @@ extern int vb12; extern int vb13; #pragma acc declare device_resident (vb13) -#pragma acc routine seq +#pragma acc routine int func5 (int a) { @@ -260,7 +302,7 @@ func5 (int a) return vb13; } -#pragma acc routine seq +#pragma acc routine int func6 (int a) { diff --git a/gcc/testsuite/g++.dg/goacc/template.C b/gcc/testsuite/g++.dg/goacc/template.C index dae92b08987..8bcd2a1ce43 100644 --- a/gcc/testsuite/g++.dg/goacc/template.C +++ b/gcc/testsuite/g++.dg/goacc/template.C @@ -1,4 +1,4 @@ -#pragma acc routine seq +#pragma acc routine template T accDouble(int val) { -- 2.30.2