[clang] 1cd981a - [OpenACC] Implement private/firstprivate for combined constructs
via cfe-commits
cfe-commits at lists.llvm.org
Thu Nov 14 09:57:45 PST 2024
Author: erichkeane
Date: 2024-11-14T09:57:39-08:00
New Revision: 1cd981a5f3c89058edd61cdeb1efa3232b1f71e6
URL: https://github.com/llvm/llvm-project/commit/1cd981a5f3c89058edd61cdeb1efa3232b1f71e6
DIFF: https://github.com/llvm/llvm-project/commit/1cd981a5f3c89058edd61cdeb1efa3232b1f71e6.diff
LOG: [OpenACC] Implement private/firstprivate for combined constructs
This is another pair of clauses where the work is already done from
previous constructs, so this just has to allow them and include tests
for them. This patch adds testing, does a few little cleanup bits on the
clause checking, and enables these.
Added:
clang/test/SemaOpenACC/combined-construct-firstprivate-clause.c
clang/test/SemaOpenACC/combined-construct-firstprivate-clause.cpp
clang/test/SemaOpenACC/combined-construct-private-clause.c
clang/test/SemaOpenACC/combined-construct-private-clause.cpp
clang/test/SemaOpenACC/combined-construct-private-firstprivate-ast.cpp
Modified:
clang/lib/Sema/SemaOpenACC.cpp
clang/test/AST/ast-print-openacc-combined-construct.cpp
clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
clang/test/SemaOpenACC/combined-construct-default-clause.c
clang/test/SemaOpenACC/compute-construct-default-clause.c
Removed:
################################################################################
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 70a0a122d82497..3554ecd7c80e81 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -598,6 +598,7 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitDefaultClause(
OpenACCClause *SemaOpenACCClauseVisitor::VisitTileClause(
SemaOpenACC::OpenACCParsedClause &Clause) {
+ // TODO OpenACC: Remove this when we get combined construct impl for this.
if (Clause.getDirectiveKind() != OpenACCDirectiveKind::Loop)
return isNotImplemented();
@@ -699,6 +700,8 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitNumGangsClause(
// Restrictions only properly implemented on 'compute' constructs, and
// 'compute' constructs are the only construct that can do anything with
// this yet, so skip/treat as unimplemented in this case.
+ // TODO OpenACC: Remove this check when we have combined constructs for this
+ // clause.
if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()))
return isNotImplemented();
@@ -754,6 +757,7 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitNumWorkersClause(
// Restrictions only properly implemented on 'compute' constructs, and
// 'compute' constructs are the only construct that can do anything with
// this yet, so skip/treat as unimplemented in this case.
+ // TODO: OpenACC: Remove when we get combined constructs.
if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()))
return isNotImplemented();
@@ -775,6 +779,7 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitVectorLengthClause(
// Restrictions only properly implemented on 'compute' constructs, and
// 'compute' constructs are the only construct that can do anything with
// this yet, so skip/treat as unimplemented in this case.
+ // TODO: OpenACC: Remove when we get combined constructs.
if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()))
return isNotImplemented();
@@ -815,14 +820,6 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitAsyncClause(
OpenACCClause *SemaOpenACCClauseVisitor::VisitPrivateClause(
SemaOpenACC::OpenACCParsedClause &Clause) {
- // Restrictions only properly implemented on 'compute' and 'loop'
- // constructs, and 'compute'/'loop' constructs are the only construct that
- // can do anything with this yet, so skip/treat as unimplemented in this
- // case.
- if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()) &&
- Clause.getDirectiveKind() != OpenACCDirectiveKind::Loop)
- return isNotImplemented();
-
// ActOnVar ensured that everything is a valid variable reference, so there
// really isn't anything to do here. GCC does some duplicate-finding, though
// it isn't apparent in the standard where this is justified.
@@ -834,12 +831,6 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitPrivateClause(
OpenACCClause *SemaOpenACCClauseVisitor::VisitFirstPrivateClause(
SemaOpenACC::OpenACCParsedClause &Clause) {
- // Restrictions only properly implemented on 'compute' constructs, and
- // 'compute' constructs are the only construct that can do anything with
- // this yet, so skip/treat as unimplemented in this case.
- if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()))
- return isNotImplemented();
-
// ActOnVar ensured that everything is a valid variable reference, so there
// really isn't anything to do here. GCC does some duplicate-finding, though
// it isn't apparent in the standard where this is justified.
@@ -1412,6 +1403,7 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitReductionClause(
// Restrictions only properly implemented on 'compute' constructs, and
// 'compute' constructs are the only construct that can do anything with
// this yet, so skip/treat as unimplemented in this case.
+ // TODO: OpenACC: Remove check once we get combined constructs for this clause.
if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()) &&
Clause.getDirectiveKind() != OpenACCDirectiveKind::Loop)
return isNotImplemented();
@@ -1502,6 +1494,7 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitReductionClause(
OpenACCClause *SemaOpenACCClauseVisitor::VisitCollapseClause(
SemaOpenACC::OpenACCParsedClause &Clause) {
+ // TODO: Remove this check once we implement this for combined constructs.
if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()) &&
Clause.getDirectiveKind() != OpenACCDirectiveKind::Loop)
return isNotImplemented();
diff --git a/clang/test/AST/ast-print-openacc-combined-construct.cpp b/clang/test/AST/ast-print-openacc-combined-construct.cpp
index f80500e7852192..a6c81d76b69ea4 100644
--- a/clang/test/AST/ast-print-openacc-combined-construct.cpp
+++ b/clang/test/AST/ast-print-openacc-combined-construct.cpp
@@ -96,4 +96,13 @@ void foo() {
// CHECK-NEXT: ;
#pragma acc serial loop default(present)
for(int i = 0;i<5;++i);
+
+// CHECK: #pragma acc parallel loop private(i, array[1], array, array[1:2])
+#pragma acc parallel loop private(i, array[1], array, array[1:2])
+ for(int i = 0;i<5;++i);
+
+// CHECK: #pragma acc serial loop firstprivate(i, array[1], array, array[1:2])
+#pragma acc serial loop firstprivate(i, array[1], array, array[1:2])
+ for(int i = 0;i<5;++i);
+
}
diff --git a/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c b/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
index 9dd0ed922ffc2c..95c9c8a47d70b7 100644
--- a/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
+++ b/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
@@ -99,8 +99,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'device_resident' not yet implemented}}
#pragma acc parallel loop auto device_resident(VarPtr)
for(unsigned i = 0; i < 5; ++i);
- // TODOexpected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'parallel loop' directive}}
- // expected-warning at +1{{OpenACC clause 'firstprivate' not yet implemented}}
#pragma acc parallel loop auto firstprivate(Var)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
@@ -117,7 +115,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'present' not yet implemented}}
#pragma acc parallel loop auto present(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'private' not yet implemented}}
#pragma acc parallel loop auto private(Var)
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'copyout' clause is not valid on 'parallel loop' directive}}
@@ -267,8 +264,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'device_resident' not yet implemented}}
#pragma acc parallel loop device_resident(VarPtr) auto
for(unsigned i = 0; i < 5; ++i);
- // TODOexpected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'parallel loop' directive}}
- // expected-warning at +1{{OpenACC clause 'firstprivate' not yet implemented}}
#pragma acc parallel loop firstprivate(Var) auto
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
@@ -285,7 +280,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'present' not yet implemented}}
#pragma acc parallel loop present(Var) auto
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'private' not yet implemented}}
#pragma acc parallel loop private(Var) auto
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'copyout' clause is not valid on 'parallel loop' directive}}
@@ -436,8 +430,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'device_resident' not yet implemented}}
#pragma acc parallel loop independent device_resident(VarPtr)
for(unsigned i = 0; i < 5; ++i);
- // TODOexpected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'parallel loop' directive}}
- // expected-warning at +1{{OpenACC clause 'firstprivate' not yet implemented}}
#pragma acc parallel loop independent firstprivate(Var)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
@@ -454,7 +446,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'present' not yet implemented}}
#pragma acc parallel loop independent present(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'private' not yet implemented}}
#pragma acc parallel loop independent private(Var)
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'copyout' clause is not valid on 'parallel loop' directive}}
@@ -604,8 +595,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'device_resident' not yet implemented}}
#pragma acc parallel loop device_resident(VarPtr) independent
for(unsigned i = 0; i < 5; ++i);
- // TODOexpected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'parallel loop' directive}}
- // expected-warning at +1{{OpenACC clause 'firstprivate' not yet implemented}}
#pragma acc parallel loop firstprivate(Var) independent
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
@@ -622,7 +611,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'present' not yet implemented}}
#pragma acc parallel loop present(Var) independent
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'private' not yet implemented}}
#pragma acc parallel loop private(Var) independent
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'copyout' clause is not valid on 'parallel loop' directive}}
@@ -779,8 +767,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'device_resident' not yet implemented}}
#pragma acc parallel loop seq device_resident(VarPtr)
for(unsigned i = 0; i < 5; ++i);
- // TODOexpected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'parallel loop' directive}}
- // expected-warning at +1{{OpenACC clause 'firstprivate' not yet implemented}}
#pragma acc parallel loop seq firstprivate(Var)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
@@ -797,7 +783,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'present' not yet implemented}}
#pragma acc parallel loop seq present(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'private' not yet implemented}}
#pragma acc parallel loop seq private(Var)
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'copyout' clause is not valid on 'parallel loop' directive}}
@@ -953,8 +938,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'device_resident' not yet implemented}}
#pragma acc parallel loop device_resident(VarPtr) seq
for(unsigned i = 0; i < 5; ++i);
- // TODOexpected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'parallel loop' directive}}
- // expected-warning at +1{{OpenACC clause 'firstprivate' not yet implemented}}
#pragma acc parallel loop firstprivate(Var) seq
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
@@ -971,7 +954,6 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'present' not yet implemented}}
#pragma acc parallel loop present(Var) seq
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'private' not yet implemented}}
#pragma acc parallel loop private(Var) seq
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'copyout' clause is not valid on 'parallel loop' directive}}
diff --git a/clang/test/SemaOpenACC/combined-construct-default-clause.c b/clang/test/SemaOpenACC/combined-construct-default-clause.c
index 058a6066c5eacb..646942ec700133 100644
--- a/clang/test/SemaOpenACC/combined-construct-default-clause.c
+++ b/clang/test/SemaOpenACC/combined-construct-default-clause.c
@@ -18,7 +18,6 @@ void SingleOnly() {
#pragma acc kernels loop self default(present) present(i) default(none) copy(i)
for(int i = 5; i < 10;++i);
- // expected-warning at +4{{OpenACC clause 'private' not yet implemented}}
// expected-warning at +3{{OpenACC clause 'copy' not yet implemented}}
// expected-error at +2{{OpenACC 'default' clause cannot appear more than once on a 'parallel loop' directive}}
// expected-note at +1{{previous clause is here}}
diff --git a/clang/test/SemaOpenACC/combined-construct-firstprivate-clause.c b/clang/test/SemaOpenACC/combined-construct-firstprivate-clause.c
new file mode 100644
index 00000000000000..107874b6d3b5aa
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-firstprivate-clause.c
@@ -0,0 +1,55 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+typedef struct IsComplete {
+ struct S { int A; } CompositeMember;
+ int ScalarMember;
+ float ArrayMember[5];
+ void *PointerMember;
+} Complete;
+void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete CompositeParam) {
+ int LocalInt;
+ short *LocalPointer;
+ float LocalArray[5];
+ Complete LocalComposite;
+ // Check Appertainment:
+#pragma acc parallel loop firstprivate(LocalInt)
+ for (int i = 5; i < 10; ++i);
+#pragma acc serial loop firstprivate(LocalInt)
+ for (int i = 5; i < 10; ++i);
+ // expected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'kernels loop' directive}}
+#pragma acc kernels loop firstprivate(LocalInt)
+ for (int i = 5; i < 10; ++i);
+
+ // Valid cases:
+#pragma acc parallel loop firstprivate(LocalInt, LocalPointer, LocalArray)
+ for (int i = 5; i < 10; ++i);
+#pragma acc serial loop firstprivate(LocalArray[2:1])
+ for (int i = 5; i < 10; ++i);
+
+#pragma acc parallel loop firstprivate(LocalComposite.ScalarMember, LocalComposite.ScalarMember)
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc serial loop firstprivate(1 + IntParam)
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc serial loop firstprivate(+IntParam)
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel loop firstprivate(PointerParam[2:])
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc serial loop firstprivate(ArrayParam[2:5])
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc parallel loop firstprivate((float*)ArrayParam[2:5])
+ for (int i = 5; i < 10; ++i);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc serial loop firstprivate((float)ArrayParam[2])
+ for (int i = 5; i < 10; ++i);
+}
diff --git a/clang/test/SemaOpenACC/combined-construct-firstprivate-clause.cpp b/clang/test/SemaOpenACC/combined-construct-firstprivate-clause.cpp
new file mode 100644
index 00000000000000..1b888ae4c6b58e
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-firstprivate-clause.cpp
@@ -0,0 +1,113 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+enum SomeE{};
+typedef struct IsComplete {
+ struct S { int A; } CompositeMember;
+ int ScalarMember;
+ float ArrayMember[5];
+ SomeE EnumMember;
+ char *PointerMember;
+} Complete;
+
+void uses(int IntParam, char *PointerParam, float ArrayParam[5], Complete CompositeParam, int &IntParamRef) {
+ int LocalInt;
+ char *LocalPointer;
+ float LocalArray[5];
+ // Check Appertainment:
+#pragma acc parallel loop firstprivate(LocalInt)
+ for (int i = 5; i < 10; ++i);
+#pragma acc serial loop firstprivate(LocalInt)
+ for (int i = 5; i < 10; ++i);
+ // expected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'kernels loop' directive}}
+#pragma acc kernels loop firstprivate(LocalInt)
+ for (int i = 5; i < 10; ++i);
+
+ // Valid cases:
+#pragma acc parallel loop firstprivate(LocalInt, LocalPointer, LocalArray)
+ for (int i = 5; i < 10; ++i);
+#pragma acc serial loop firstprivate(LocalArray[2:1])
+ for (int i = 5; i < 10; ++i);
+
+ Complete LocalComposite2;
+#pragma acc parallel loop firstprivate(LocalComposite2.ScalarMember, LocalComposite2.ScalarMember)
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc parallel loop firstprivate(1 + IntParam)
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc serial loop firstprivate(+IntParam)
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel loop firstprivate(PointerParam[2:])
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel loop firstprivate(ArrayParam[2:5])
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc serial loop firstprivate((float*)ArrayParam[2:5])
+ for (int i = 5; i < 10; ++i);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc parallel loop firstprivate((float)ArrayParam[2])
+ for (int i = 5; i < 10; ++i);
+}
+
+template<typename T, unsigned I, typename V>
+void TemplUses(T t, T (&arrayT)[I], V TemplComp) {
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc parallel loop firstprivate(+t)
+ for (int i = 5; i < 10; ++i);
+
+ // NTTP's are only valid if it is a reference to something.
+ // expected-error at +2{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+ // expected-note@#TEMPL_USES_INST{{in instantiation of}}
+#pragma acc parallel loop firstprivate(I)
+ for (int i = 5; i < 10; ++i);
+
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc serial loop firstprivate(t, I)
+ for (int i = 5; i < 10; ++i);
+
+#pragma acc parallel loop firstprivate(arrayT)
+ for (int i = 5; i < 10; ++i);
+
+#pragma acc parallel loop firstprivate(TemplComp)
+ for (int i = 5; i < 10; ++i);
+
+#pragma acc parallel loop firstprivate(TemplComp.PointerMember[5])
+ for (int i = 5; i < 10; ++i);
+ int *Pointer;
+#pragma acc serial loop firstprivate(Pointer[:I])
+ for (int i = 5; i < 10; ++i);
+#pragma acc parallel loop firstprivate(Pointer[:t])
+ for (int i = 5; i < 10; ++i);
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel loop firstprivate(Pointer[1:])
+ for (int i = 5; i < 10; ++i);
+}
+
+template<unsigned I, auto &NTTP_REF>
+void NTTP() {
+ // NTTP's are only valid if it is a reference to something.
+ // expected-error at +2{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+ // expected-note@#NTTP_INST{{in instantiation of}}
+#pragma acc parallel loop firstprivate(I)
+ for (int i = 5; i < 10; ++i);
+
+#pragma acc serial loop firstprivate(NTTP_REF)
+ for (int i = 5; i < 10; ++i);
+}
+
+void Inst() {
+ static constexpr int NTTP_REFed = 1;
+ int i;
+ int Arr[5];
+ Complete C;
+ TemplUses(i, Arr, C); // #TEMPL_USES_INST
+ NTTP<5, NTTP_REFed>(); // #NTTP_INST
+}
diff --git a/clang/test/SemaOpenACC/combined-construct-private-clause.c b/clang/test/SemaOpenACC/combined-construct-private-clause.c
new file mode 100644
index 00000000000000..1cd30db8f4aa79
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-private-clause.c
@@ -0,0 +1,128 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+struct Incomplete;
+enum SomeE{ A };
+typedef struct IsComplete {
+ struct S { int A; } CompositeMember;
+ int ScalarMember;
+ float ArrayMember[5];
+ enum SomeE EnumMember;
+ void *PointerMember;
+} Complete;
+
+int GlobalInt;
+float GlobalArray[5];
+short *GlobalPointer;
+Complete GlobalComposite;
+
+void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete CompositeParam) {
+ int LocalInt;
+ short *LocalPointer;
+ float LocalArray[5];
+ Complete LocalComposite;
+
+ // Valid cases:
+#pragma acc parallel loop private(LocalInt, LocalPointer, LocalArray)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(LocalArray)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(LocalArray[:])
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(LocalArray[:5])
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(LocalArray[2:])
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(LocalArray[2:1])
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(LocalArray[2])
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(LocalComposite)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(LocalComposite.EnumMember)
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(LocalComposite.ScalarMember)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(LocalComposite.ArrayMember)
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(LocalComposite.ArrayMember[5])
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(LocalComposite.PointerMember)
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(GlobalInt, GlobalArray, GlobalPointer, GlobalComposite)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(GlobalArray[2], GlobalPointer[2], GlobalComposite.CompositeMember.A)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(LocalComposite, GlobalComposite)
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(IntParam, PointerParam, ArrayParam, CompositeParam)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(PointerParam[IntParam], ArrayParam[IntParam], CompositeParam.CompositeMember.A)
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc parallel loop private(LocalArray) private(LocalArray[2])
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc serial loop private(LocalArray, LocalArray[2])
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc kernels loop private(LocalComposite, LocalComposite.ScalarMember)
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc parallel loop private(LocalComposite.CompositeMember.A, LocalComposite.ScalarMember)
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc serial loop private(LocalComposite.CompositeMember.A) private(LocalComposite.ScalarMember)
+ for(int i = 0; i < 5; ++i);
+
+ Complete LocalComposite2;
+#pragma acc kernels loop private(LocalComposite2.ScalarMember, LocalComposite2.ScalarMember)
+ for(int i = 0; i < 5; ++i);
+
+ // Invalid cases, arbitrary expressions.
+ struct Incomplete *I;
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc parallel loop private(*I)
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc serial loop private(GlobalInt + IntParam)
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc kernels loop private(+GlobalInt)
+ for(int i = 0; i < 5; ++i);
+
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel loop private(PointerParam[:])
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(PointerParam[:5])
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(PointerParam[:IntParam])
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel loop private(PointerParam[2:])
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(PointerParam[2:5])
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(PointerParam[2])
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(ArrayParam[:])
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(ArrayParam[:5])
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(ArrayParam[:IntParam])
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(ArrayParam[2:])
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc serial loop private(ArrayParam[2:5])
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(ArrayParam[2])
+ for(int i = 0; i < 5; ++i);
+
+ // expected-error at +2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc parallel loop private((float*)ArrayParam[2:5])
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc kernels loop private((float)ArrayParam[2])
+ for(int i = 0; i < 5; ++i);
+}
diff --git a/clang/test/SemaOpenACC/combined-construct-private-clause.cpp b/clang/test/SemaOpenACC/combined-construct-private-clause.cpp
new file mode 100644
index 00000000000000..d94c336c21d97a
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-private-clause.cpp
@@ -0,0 +1,150 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+struct Incomplete;
+enum SomeE{};
+typedef struct IsComplete {
+ struct S { int A; } CompositeMember;
+ int ScalarMember;
+ float ArrayMember[5];
+ SomeE EnumMember;
+ char *PointerMember;
+} Complete;
+
+int GlobalInt;
+float GlobalArray[5];
+char *GlobalPointer;
+Complete GlobalComposite;
+
+void uses(int IntParam, char *PointerParam, float ArrayParam[5], Complete CompositeParam, int &IntParamRef) {
+ int LocalInt;
+ char *LocalPointer;
+ float LocalArray[5];
+ Complete LocalComposite;
+
+ // Valid cases:
+#pragma acc parallel loop private(LocalInt, LocalPointer, LocalArray)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(LocalArray)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(LocalArray[2])
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(LocalComposite)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(LocalComposite.EnumMember)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(LocalComposite.ScalarMember)
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(LocalComposite.ArrayMember)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(LocalComposite.ArrayMember[5])
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(LocalComposite.PointerMember)
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(GlobalInt, GlobalArray, GlobalPointer, GlobalComposite)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(GlobalArray[2], GlobalPointer[2], GlobalComposite.CompositeMember.A)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(LocalComposite, GlobalComposite)
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop private(IntParam, PointerParam, ArrayParam, CompositeParam) private(IntParamRef)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop private(PointerParam[IntParam], ArrayParam[IntParam], CompositeParam.CompositeMember.A)
+ for(int i = 0; i < 5; ++i);
+
+
+ // Invalid cases, arbitrary expressions.
+ Incomplete *I;
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc kernels loop private(*I)
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc parallel loop private(GlobalInt + IntParam)
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc serial loop private(+GlobalInt)
+ for(int i = 0; i < 5; ++i);
+}
+
+template<typename T, unsigned I, typename V>
+void TemplUses(T t, T (&arrayT)[I], V TemplComp) {
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc kernels loop private(+t)
+ for(int i = 0; i < 5; ++i);
+
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc parallel loop private(+I)
+ for(int i = 0; i < 5; ++i);
+
+ // NTTP's are only valid if it is a reference to something.
+ // expected-error at +2{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+ // expected-note@#TEMPL_USES_INST{{in instantiation of}}
+#pragma acc serial loop private(I)
+ for(int i = 0; i < 5; ++i);
+
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+#pragma acc kernels loop private(t, I)
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc parallel loop private(arrayT)
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc serial loop private(TemplComp)
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc kernels loop private(TemplComp.PointerMember[5])
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc parallel loop private(TemplComp.PointerMember[5]) private(TemplComp)
+ for(int i = 0; i < 5; ++i);
+
+ int *Pointer;
+#pragma acc serial loop private(Pointer[:I])
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop private(Pointer[:t])
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel loop private(Pointer[1:])
+ for(int i = 0; i < 5; ++i);
+}
+
+template<unsigned I, auto &NTTP_REF>
+void NTTP() {
+ // NTTP's are only valid if it is a reference to something.
+ // expected-error at +2{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
+ // expected-note@#NTTP_INST{{in instantiation of}}
+#pragma acc serial loop private(I)
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc kernels loop private(NTTP_REF)
+ for(int i = 0; i < 5; ++i);
+}
+
+struct S {
+ int ThisMember;
+ int ThisMemberArray[5];
+
+ void foo();
+};
+
+void S::foo() {
+#pragma acc parallel loop private(ThisMember, this->ThisMemberArray[1])
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc serial loop private(ThisMemberArray[1:2])
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc kernels loop private(this)
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc parallel loop private(ThisMember, this->ThisMember)
+ for(int i = 0; i < 5; ++i);
+}
+
+void Inst() {
+ static constexpr int NTTP_REFed = 1;
+ int i;
+ int Arr[5];
+ Complete C;
+ TemplUses(i, Arr, C); // #TEMPL_USES_INST
+ NTTP<5, NTTP_REFed>(); // #NTTP_INST
+}
diff --git a/clang/test/SemaOpenACC/combined-construct-private-firstprivate-ast.cpp b/clang/test/SemaOpenACC/combined-construct-private-firstprivate-ast.cpp
new file mode 100644
index 00000000000000..03c258aca4f6aa
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-private-firstprivate-ast.cpp
@@ -0,0 +1,226 @@
+// RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s
+
+// Test this with PCH.
+// RUN: %clang_cc1 %s -fopenacc -emit-pch -o %t %s
+// RUN: %clang_cc1 %s -fopenacc -include-pch %t -ast-dump-all | FileCheck %s
+#ifndef PCH_HELPER
+#define PCH_HELPER
+
+int Global;
+short GlobalArray[5];
+
+void NormalUses(float *PointerParam) {
+ // CHECK: FunctionDecl{{.*}}NormalUses
+ // CHECK: ParmVarDecl
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel loop private(Global, GlobalArray[2])
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'short' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'short *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 2
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+#pragma acc serial loop private(GlobalArray, PointerParam[Global])
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'float' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+#pragma acc parallel loop firstprivate(GlobalArray, PointerParam[Global])
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: firstprivate clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'float' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+#pragma acc serial loop firstprivate(GlobalArray, PointerParam[Global : Global])
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
+ // CHECK-NEXT: firstprivate clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}} 'PointerParam' 'float *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+}
+template<auto &NTTP, typename T, typename U>
+void TemplUses(T t, U u, T*PointerParam) {
+
+ // CHECK: FunctionTemplateDecl{{.*}}TemplUses
+ // CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}referenced 'auto &' depth 0 index 0 NTTP
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 1 T
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 2 U
+ // CHECK-NEXT: FunctionDecl{{.*}} TemplUses 'void (T, U, T *)'
+ // CHECK-NEXT: ParmVarDecl{{.*}} referenced t 'T'
+ // CHECK-NEXT: ParmVarDecl{{.*}} referenced u 'U'
+ // CHECK-NEXT: ParmVarDecl{{.*}} referenced PointerParam 'T *'
+ // CHECK-NEXT: CompoundStmt
+#pragma acc serial loop private(GlobalArray, PointerParam[Global])
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'T' lvalue
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T *' lvalue ParmVar{{.*}}'PointerParam' 'T *'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+#pragma acc kernels loop private(t, u)
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+#pragma acc parallel loop firstprivate(t, u)
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: firstprivate clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+#pragma acc kernels loop private(t) private(u)
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+#pragma acc kernels loop private(t) private(NTTP, u)
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+#pragma acc parallel loop private(t) firstprivate(NTTP, u)
+ for (int i = 5; i < 10; ++i);
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: firstprivate clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+ // Check the instantiated versions of the above.
+ // CHECK-NEXT: FunctionDecl{{.*}} used TemplUses 'void (int, int *, int *)' implicit_instantiation
+ // CHECK-NEXT: TemplateArgument decl
+ // CHECK-NEXT: Var{{.*}} 'CEVar' 'const unsigned int'
+ // CHECK-NEXT: TemplateArgument type 'int'
+ // CHECK-NEXT: BuiltinType{{.*}} 'int'
+ // CHECK-NEXT: TemplateArgument type 'int[1]'
+ // CHECK-NEXT: ConstantArrayType{{.*}} 'int[1]'{{.*}} 1
+ // CHECK-NEXT: BuiltinType{{.*}} 'int'
+ // CHECK-NEXT: ParmVarDecl{{.*}} used t 'int'
+ // CHECK-NEXT: ParmVarDecl{{.*}} used u 'int *'
+ // CHECK-NEXT: ParmVarDecl{{.*}} used PointerParam 'int *'
+ // CHECK-NEXT: CompoundStmt
+
+// #pragma acc serial loop private(GlobalArray, PointerParam[Global])
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'int' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}}'PointerParam' 'int *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+// #pragma acc kernels loop private(t, u)
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+// #pragma acc parallel loop firstprivate(t, u)
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: firstprivate clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+// #pragma acc kernels loop private(t) private(u)
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+// #pragma acc kernels loop private(t) private(NTTP, u)
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}}'const unsigned int' lvalue
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'auto &' depth 0 index 0 NTTP
+ // CHECK-NEXT: DeclRefExpr{{.*}}'const unsigned int' lvalue Var{{.*}} 'CEVar' 'const unsigned int'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+
+// #pragma acc parallel loop private(t) firstprivate(NTTP, u)
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: firstprivate clause
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}}'const unsigned int' lvalue
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'auto &' depth 0 index 0 NTTP
+ // CHECK-NEXT: DeclRefExpr{{.*}}'const unsigned int' lvalue Var{{.*}} 'CEVar' 'const unsigned int'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: ForStmt
+ // CHECK: NullStmt
+}
+
+void Inst() {
+ static constexpr unsigned CEVar = 1;
+ int i;
+ int Arr[5];
+ TemplUses<CEVar, int, int[1]>({}, {}, &i);
+}
+#endif // PCH_HELPER
diff --git a/clang/test/SemaOpenACC/compute-construct-default-clause.c b/clang/test/SemaOpenACC/compute-construct-default-clause.c
index 15d6c512d6e654..e71fee47e8a292 100644
--- a/clang/test/SemaOpenACC/compute-construct-default-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-default-clause.c
@@ -16,7 +16,6 @@ void SingleOnly() {
#pragma acc kernels self default(present) present(i) default(none) copy(i)
while(0);
- // expected-warning at +4{{OpenACC clause 'private' not yet implemented}}
// expected-warning at +3{{OpenACC clause 'copy' not yet implemented}}
// expected-error at +2{{OpenACC 'default' clause cannot appear more than once on a 'parallel loop' directive}}
// expected-note at +1{{previous clause is here}}
More information about the cfe-commits
mailing list