[clang] 6b2de10 - [OpenACC] implement 'device_type' for combined constructs
via cfe-commits
cfe-commits at lists.llvm.org
Wed Nov 13 09:57:46 PST 2024
Author: erichkeane
Date: 2024-11-13T09:57:40-08:00
New Revision: 6b2de10c687dedb8e460699d2b68f0b0eafc2b4e
URL: https://github.com/llvm/llvm-project/commit/6b2de10c687dedb8e460699d2b68f0b0eafc2b4e
DIFF: https://github.com/llvm/llvm-project/commit/6b2de10c687dedb8e460699d2b68f0b0eafc2b4e.diff
LOG: [OpenACC] implement 'device_type' for combined constructs
This clause is pretty small/doesn't do much semantic-analysis-wise, , other than
have two spellings and disallow certain clauses after it. However, as
most of those aren't implemented yet, the diagnostic is left as a TODO.
Added:
clang/test/SemaOpenACC/combined-construct-device_type-ast.cpp
clang/test/SemaOpenACC/combined-construct-device_type-clause.c
clang/test/SemaOpenACC/combined-construct-device_type-clause.cpp
Modified:
clang/include/clang/Basic/DiagnosticSemaKinds.td
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.cpp
clang/test/SemaOpenACC/compute-construct-device_type-clause.c
clang/test/SemaOpenACC/compute-construct-num_gangs-clause.cpp
clang/test/SemaOpenACC/loop-construct.cpp
Removed:
################################################################################
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index dc9f36c308db91..17eb28e8fc5623 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12686,7 +12686,7 @@ def err_acc_var_not_pointer_type
def note_acc_expected_pointer_var : Note<"expected variable of pointer type">;
def err_acc_clause_after_device_type
: Error<"OpenACC clause '%0' may not follow a '%1' clause in a "
- "%select{'%3'|compute}2 construct">;
+ "'%2' construct">;
def err_acc_clause_cannot_combine
: Error<"OpenACC clause '%0' may not appear on the same construct as a "
"'%1' clause on a '%2' construct">;
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index d33c84de21709c..7bf99eb7148761 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -502,7 +502,6 @@ bool checkValidAfterDeviceType(
}
S.Diag(NewClause.getBeginLoc(), diag::err_acc_clause_after_device_type)
<< NewClause.getClauseKind() << DeviceTypeClause.getClauseKind()
- << isOpenACCComputeDirectiveKind(NewClause.getDirectiveKind())
<< NewClause.getDirectiveKind();
S.Diag(DeviceTypeClause.getBeginLoc(), diag::note_acc_previous_clause_here);
return true;
@@ -999,12 +998,13 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitWaitClause(
OpenACCClause *SemaOpenACCClauseVisitor::VisitDeviceTypeClause(
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.
+ // Restrictions only properly implemented on 'compute', 'combined', and
+ // 'loop' constructs, and 'compute'/'combined'/'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)
+ Clause.getDirectiveKind() != OpenACCDirectiveKind::Loop &&
+ !isOpenACCCombinedDirectiveKind(Clause.getDirectiveKind()))
return isNotImplemented();
// TODO OpenACC: Once we get enough of the CodeGen implemented that we have
diff --git a/clang/test/AST/ast-print-openacc-combined-construct.cpp b/clang/test/AST/ast-print-openacc-combined-construct.cpp
index 5318640eddf535..d77465c69f4acb 100644
--- a/clang/test/AST/ast-print-openacc-combined-construct.cpp
+++ b/clang/test/AST/ast-print-openacc-combined-construct.cpp
@@ -31,4 +31,45 @@ void foo() {
// CHECK-NEXT: ;
#pragma acc kernels loop independent
for(int i = 0;i<5;++i);
+
+ bool SomeB;
+ struct SomeStruct{} SomeStructImpl;
+
+//CHECK: #pragma acc parallel loop dtype(SomeB)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc parallel loop dtype(SomeB)
+ for(int i = 0;i<5;++i);
+
+//CHECK: #pragma acc serial loop device_type(SomeStruct)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc serial loop device_type(SomeStruct)
+ for(int i = 0;i<5;++i);
+
+//CHECK: #pragma acc kernels loop device_type(int)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc kernels loop device_type(int)
+ for(int i = 0;i<5;++i);
+
+//CHECK: #pragma acc parallel loop dtype(bool)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc parallel loop dtype(bool)
+ for(int i = 0;i<5;++i);
+
+//CHECK: #pragma acc serial loop device_type(SomeStructImpl)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc serial loop device_type (SomeStructImpl)
+ for(int i = 0;i<5;++i);
+
+// CHECK: #pragma acc kernels loop dtype(AnotherIdent)
+// CHECK-NEXT: for (int i = 0; i < 5; ++i)
+// CHECK-NEXT: ;
+#pragma acc kernels loop dtype(AnotherIdent)
+ 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 d776882b768448..95cc7a7ed22949 100644
--- a/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
+++ b/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
@@ -195,10 +195,8 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'default_async' not yet implemented}}
#pragma acc parallel loop auto default_async(1)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device_type' not yet implemented}}
#pragma acc parallel loop auto device_type(*)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'dtype' not yet implemented}}
#pragma acc parallel loop auto dtype(*)
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'async' clause is not valid on 'parallel loop' directive}}
@@ -371,10 +369,8 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'default_async' not yet implemented}}
#pragma acc parallel loop default_async(1) auto
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device_type' not yet implemented}}
#pragma acc parallel loop device_type(*) auto
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'dtype' not yet implemented}}
#pragma acc parallel loop dtype(*) auto
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'async' clause is not valid on 'parallel loop' directive}}
@@ -548,10 +544,8 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'default_async' not yet implemented}}
#pragma acc parallel loop independent default_async(1)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device_type' not yet implemented}}
#pragma acc parallel loop independent device_type(*)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'dtype' not yet implemented}}
#pragma acc parallel loop independent dtype(*)
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'async' clause is not valid on 'parallel loop' directive}}
@@ -724,10 +718,8 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'default_async' not yet implemented}}
#pragma acc parallel loop default_async(1) independent
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device_type' not yet implemented}}
#pragma acc parallel loop device_type(*) independent
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'dtype' not yet implemented}}
#pragma acc parallel loop dtype(*) independent
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'async' clause is not valid on 'parallel loop' directive}}
@@ -907,10 +899,8 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'default_async' not yet implemented}}
#pragma acc parallel loop seq default_async(1)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device_type' not yet implemented}}
#pragma acc parallel loop seq device_type(*)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'dtype' not yet implemented}}
#pragma acc parallel loop seq dtype(*)
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'async' clause is not valid on 'parallel loop' directive}}
@@ -1089,10 +1079,8 @@ void uses() {
// expected-warning at +1{{OpenACC clause 'default_async' not yet implemented}}
#pragma acc parallel loop default_async(1) seq
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device_type' not yet implemented}}
#pragma acc parallel loop device_type(*) seq
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'dtype' not yet implemented}}
#pragma acc parallel loop dtype(*) seq
for(unsigned i = 0; i < 5; ++i);
// TODOexpected-error at +1{{OpenACC 'async' clause is not valid on 'parallel loop' directive}}
diff --git a/clang/test/SemaOpenACC/combined-construct-device_type-ast.cpp b/clang/test/SemaOpenACC/combined-construct-device_type-ast.cpp
new file mode 100644
index 00000000000000..abc65dfc9e5227
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-device_type-ast.cpp
@@ -0,0 +1,178 @@
+// 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
+
+struct SomeS{};
+void NormalUses() {
+ // CHECK: FunctionDecl{{.*}}NormalUses
+ // CHECK-NEXT: CompoundStmt
+
+ SomeS SomeImpl;
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} SomeImpl 'SomeS'
+ // CHECK-NEXT: CXXConstructExpr
+ bool SomeVar;
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} SomeVar 'bool'
+
+#pragma acc parallel loop device_type(SomeS) dtype(SomeImpl)
+ for(int i = 0; i < 5; ++i){}
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: device_type(SomeS)
+ // CHECK-NEXT: dtype(SomeImpl)
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: CompoundStmt
+#pragma acc serial loop device_type(SomeVar) dtype(int)
+ for(int i = 0; i < 5; ++i){}
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
+ // CHECK-NEXT: device_type(SomeVar)
+ // CHECK-NEXT: dtype(int)
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: CompoundStmt
+#pragma acc kernels loop device_type(private) dtype(struct)
+ for(int i = 0; i < 5; ++i){}
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: device_type(private)
+ // CHECK-NEXT: dtype(struct)
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: CompoundStmt
+#pragma acc parallel loop device_type(private) dtype(class)
+ for(int i = 0; i < 5; ++i){}
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: device_type(private)
+ // CHECK-NEXT: dtype(class)
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: CompoundStmt
+#pragma acc serial loop device_type(float) dtype(*)
+ for(int i = 0; i < 5; ++i){}
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} serial loop
+ // CHECK-NEXT: device_type(float)
+ // CHECK-NEXT: dtype(*)
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: CompoundStmt
+#pragma acc kernels loop device_type(float, int) dtype(*)
+ for(int i = 0; i < 5; ++i){}
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} kernels loop
+ // CHECK-NEXT: device_type(float, int)
+ // CHECK-NEXT: dtype(*)
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: CompoundStmt
+}
+
+template<typename T>
+void TemplUses() {
+ // CHECK-NEXT: FunctionTemplateDecl{{.*}}TemplUses
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}}T
+ // CHECK-NEXT: FunctionDecl{{.*}}TemplUses
+ // CHECK-NEXT: CompoundStmt
+#pragma acc parallel loop device_type(T) dtype(T)
+ for(int i = 0; i < 5; ++i){}
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: device_type(T)
+ // CHECK-NEXT: dtype(T)
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: CompoundStmt
+
+
+ // Instantiations
+ // CHECK-NEXT: FunctionDecl{{.*}} TemplUses 'void ()' implicit_instantiation
+ // CHECK-NEXT: TemplateArgument type 'int'
+ // CHECK-NEXT: BuiltinType{{.*}} 'int'
+ // CHECK-NEXT: CompoundStmt
+
+ // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
+ // CHECK-NEXT: device_type(T)
+ // CHECK-NEXT: dtype(T)
+ // CHECK-NEXT: ForStmt
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} i 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: BinaryOperator{{.*}}'<'
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 5
+ // CHECK-NEXT: UnaryOperator{{.*}}++
+ // CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
+ // CHECK-NEXT: CompoundStmt
+}
+
+void Inst() {
+ TemplUses<int>();
+}
+
+#endif // PCH_HELPER
diff --git a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
new file mode 100644
index 00000000000000..3d5cb7eb305b03
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
@@ -0,0 +1,228 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+#define MACRO +FOO
+
+void uses() {
+ typedef struct S{} STy;
+ STy SImpl;
+
+#pragma acc parallel loop device_type(I)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop device_type(S) dtype(STy)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop dtype(SImpl)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop dtype(int) device_type(*)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop dtype(true) device_type(false)
+ for(int i = 0; i < 5; ++i);
+
+ // expected-error at +1{{expected identifier}}
+#pragma acc kernels loop dtype(int, *)
+ for(int i = 0; i < 5; ++i);
+
+#pragma acc parallel loop device_type(I, int)
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +2{{expected ','}}
+ // expected-error at +1{{expected identifier}}
+#pragma acc kernels loop dtype(int{})
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{expected identifier}}
+#pragma acc kernels loop dtype(5)
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{expected identifier}}
+#pragma acc kernels loop dtype(MACRO)
+ for(int i = 0; i < 5; ++i);
+
+ // Compute constructs allow 'async', 'wait', num_gangs', 'num_workers',
+ // 'vector_length' after 'device_type', loop allows 'collapse', 'gang',
+ // 'worker', 'vector', 'seq', 'independent', 'auto', and 'tile' after
+ // 'device_type'.
+
+ //expected-warning at +1{{OpenACC clause 'vector' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) vector
+ for(int i = 0; i < 5; ++i);
+
+ // TODOexpected-error at +2{{OpenACC clause 'finalize' may not follow a 'device_type' clause in a 'serial loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) finalize
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'if_present' may not follow a 'device_type' clause in a 'kernels loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'if_present' not yet implemented, clause ignored}}
+#pragma acc kernels loop device_type(*) if_present
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop device_type(*) seq
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop device_type(*) independent
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop device_type(*) auto
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'worker' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) worker
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'nohost' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'nohost' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) nohost
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'default' not yet implemented, clause ignored}}
+#pragma acc kernels loop device_type(*) default(none)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'if' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) if(1)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'self' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) self
+ for(int i = 0; i < 5; ++i);
+
+ int Var;
+ int *VarPtr;
+ // expected-warning at +1{{OpenACC clause 'copy' not yet implemented, clause ignored}}
+#pragma acc kernels loop device_type(*) copy(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +2{{OpenACC clause name 'pcopy' is a deprecated clause name and is now an alias for 'copy'}}
+ // expected-warning at +1{{OpenACC clause 'pcopy' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) pcopy(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +2{{OpenACC clause name 'present_or_copy' is a deprecated clause name and is now an alias for 'copy'}}
+ // expected-warning at +1{{OpenACC clause 'present_or_copy' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) present_or_copy(Var)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'use_device' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented, clause ignored}}
+#pragma acc kernels loop device_type(*) use_device(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'attach' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) attach(Var)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'delete' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'delete' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) delete(Var)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'detach' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'detach' not yet implemented, clause ignored}}
+#pragma acc kernels loop device_type(*) detach(Var)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'device' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'device' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) device(VarPtr)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'deviceptr' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) deviceptr(VarPtr)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'device_resident' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'device_resident' not yet implemented, clause ignored}}
+#pragma acc kernels loop device_type(*) device_resident(VarPtr)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'firstprivate' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) firstprivate(Var)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'host' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'host' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) host(Var)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'link' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'link' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) link(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'no_create' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) no_create(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'present' not yet implemented, clause ignored}}
+#pragma acc kernels loop device_type(*) present(Var)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'private' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'private' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) private(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) copyout(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +2{{OpenACC clause name 'pcopyout' is a deprecated clause name and is now an alias for 'copyout'}}
+ // expected-warning at +1{{OpenACC clause 'pcopyout' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) pcopyout(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +2{{OpenACC clause name 'present_or_copyout' is a deprecated clause name and is now an alias for 'copyout'}}
+ // expected-warning at +1{{OpenACC clause 'present_or_copyout' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) present_or_copyout(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) copyin(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +2{{OpenACC clause name 'pcopyin' is a deprecated clause name and is now an alias for 'copyin'}}
+ // expected-warning at +1{{OpenACC clause 'pcopyin' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) pcopyin(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +2{{OpenACC clause name 'present_or_copyin' is a deprecated clause name and is now an alias for 'copyin'}}
+ // expected-warning at +1{{OpenACC clause 'present_or_copyin' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) present_or_copyin(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'create' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) create(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +2{{OpenACC clause name 'pcreate' is a deprecated clause name and is now an alias for 'create'}}
+ // expected-warning at +1{{OpenACC clause 'pcreate' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) pcreate(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +2{{OpenACC clause name 'present_or_create' is a deprecated clause name and is now an alias for 'create'}}
+ // expected-warning at +1{{OpenACC clause 'present_or_create' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) present_or_create(Var)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'reduction' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'reduction' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) reduction(+:Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) collapse(1)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'bind' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'bind' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) bind(Var)
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC 'vector_length' clause is not valid on 'serial loop' directive}}
+#pragma acc serial loop device_type(*) vector_length(1)
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +1{{OpenACC 'num_gangs' clause is not valid on 'serial loop' directive}}
+#pragma acc serial loop device_type(*) num_gangs(1)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'num_workers' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) num_workers(1)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'device_num' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'device_num' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) device_num(1)
+ for(int i = 0; i < 5; ++i);
+ // TODOexpected-error at +2{{OpenACC clause 'default_async' may not follow a 'device_type' clause in a 'loop' construct}}
+ // TODOexpected-note at +1{{previous clause is here}}
+ // expected-warning at +1{{OpenACC clause 'default_async' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) default_async(1)
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'async' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) async
+ for(int i = 0; i < 5; ++i);
+
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
+#pragma acc serial loop device_type(*) tile(*, 1)
+ for(int j = 0; j < 5; ++j)
+ for(int i = 0; i < 5; ++i);
+
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
+#pragma acc serial loop dtype(*) gang
+ for(int i = 0; i < 5; ++i);
+ // expected-warning at +1{{OpenACC clause 'wait' not yet implemented, clause ignored}}
+#pragma acc parallel loop device_type(*) wait
+ for(int i = 0; i < 5; ++i);
+}
diff --git a/clang/test/SemaOpenACC/combined-construct-device_type-clause.cpp b/clang/test/SemaOpenACC/combined-construct-device_type-clause.cpp
new file mode 100644
index 00000000000000..b46709aa8d4ce3
--- /dev/null
+++ b/clang/test/SemaOpenACC/combined-construct-device_type-clause.cpp
@@ -0,0 +1,25 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+template<typename T>
+void TemplUses() {
+#pragma acc parallel loop device_type(I)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop dtype(*)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop device_type(class)
+ for(int i = 0; i < 5; ++i);
+#pragma acc parallel loop device_type(private)
+ for(int i = 0; i < 5; ++i);
+#pragma acc serial loop device_type(bool)
+ for(int i = 0; i < 5; ++i);
+#pragma acc kernels loop dtype(true) device_type(false)
+ for(int i = 0; i < 5; ++i);
+ // expected-error at +2{{expected ','}}
+ // expected-error at +1{{expected identifier}}
+#pragma acc kernels loop device_type(T::value)
+ for(int i = 0; i < 5; ++i);
+}
+
+void Inst() {
+ TemplUses<int>(); // #INST
+}
diff --git a/clang/test/SemaOpenACC/combined-construct.cpp b/clang/test/SemaOpenACC/combined-construct.cpp
index 75fcbbd851f4aa..b0fd05e5a9a183 100644
--- a/clang/test/SemaOpenACC/combined-construct.cpp
+++ b/clang/test/SemaOpenACC/combined-construct.cpp
@@ -28,7 +28,6 @@ struct SomeRAIterator {
int operator*();
void operator+=(int);
bool operator!=(SomeRAIterator&);
- // TODO
};
struct HasIteratorCollection {
diff --git a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
index aaf8b76e1f3dfb..0ae972d2a99ff4 100644
--- a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
@@ -44,11 +44,11 @@ void uses() {
// Only 'async', 'wait', num_gangs', 'num_workers', 'vector_length' allowed after 'device_type'.
- // expected-error at +2{{OpenACC clause 'finalize' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'finalize' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) finalize
while(1);
- // expected-error at +2{{OpenACC clause 'if_present' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'if_present' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) if_present
while(1);
@@ -64,133 +64,133 @@ void uses() {
// expected-error at +1{{OpenACC 'worker' clause is not valid on 'kernels' directive}}
#pragma acc kernels device_type(*) worker
while(1);
- // expected-error at +2{{OpenACC clause 'nohost' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'nohost' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) nohost
while(1);
- // expected-error at +2{{OpenACC clause 'default' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'default' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) default(none)
while(1);
- // expected-error at +2{{OpenACC clause 'if' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'if' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) if(1)
while(1);
- // expected-error at +2{{OpenACC clause 'self' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'self' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) self
while(1);
int Var;
int *VarPtr;
- // expected-error at +2{{OpenACC clause 'copy' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'copy' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) copy(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'pcopy' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'pcopy' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) pcopy(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'present_or_copy' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'present_or_copy' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) present_or_copy(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'use_device' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'use_device' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) use_device(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'attach' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'attach' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) attach(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'delete' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'delete' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) delete(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'detach' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'detach' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) detach(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'device' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'device' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) device(VarPtr)
while(1);
- // expected-error at +2{{OpenACC clause 'deviceptr' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'deviceptr' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) deviceptr(VarPtr)
while(1);
- // expected-error at +2{{OpenACC clause 'device_resident' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'device_resident' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) device_resident(VarPtr)
while(1);
- // expected-error at +2{{OpenACC clause 'firstprivate' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'firstprivate' may not follow a 'device_type' clause in a 'parallel' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc parallel device_type(*) firstprivate(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'host' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'host' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) host(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'link' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'link' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) link(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'no_create' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'no_create' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) no_create(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'present' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'present' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) present(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'private' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'private' may not follow a 'device_type' clause in a 'parallel' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc parallel device_type(*) private(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'copyout' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'copyout' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) copyout(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'pcopyout' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'pcopyout' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) pcopyout(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'present_or_copyout' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'present_or_copyout' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) present_or_copyout(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'copyin' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'copyin' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) copyin(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'pcopyin' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'pcopyin' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) pcopyin(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'present_or_copyin' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'present_or_copyin' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) present_or_copyin(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'create' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'create' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) create(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'pcreate' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'pcreate' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) pcreate(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'present_or_create' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'present_or_create' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) present_or_create(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'reduction' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'reduction' may not follow a 'device_type' clause in a 'serial' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc serial device_type(*) reduction(+:Var)
while(1);
// expected-error at +1{{OpenACC 'collapse' clause is not valid on 'kernels' directive}}
#pragma acc kernels device_type(*) collapse(1)
while(1);
- // expected-error at +2{{OpenACC clause 'bind' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'bind' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) bind(Var)
while(1);
@@ -200,11 +200,11 @@ void uses() {
while(1);
#pragma acc kernels device_type(*) num_workers(1)
while(1);
- // expected-error at +2{{OpenACC clause 'device_num' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'device_num' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) device_num(1)
while(1);
- // expected-error at +2{{OpenACC clause 'default_async' may not follow a 'device_type' clause in a compute construct}}
+ // expected-error at +2{{OpenACC clause 'default_async' may not follow a 'device_type' clause in a 'kernels' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc kernels device_type(*) default_async(1)
while(1);
diff --git a/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.cpp b/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.cpp
index ec3df87a065572..c50f52afda7c1c 100644
--- a/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.cpp
+++ b/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.cpp
@@ -105,7 +105,6 @@ void Test() {
// expected-error at +1{{OpenACC 'num_gangs' clause is not valid on 'serial' directive}}
#pragma acc serial num_gangs(Explicit, NC, Ambiguous)
while(1);
- // TODO
}
struct HasInt {
diff --git a/clang/test/SemaOpenACC/loop-construct.cpp b/clang/test/SemaOpenACC/loop-construct.cpp
index 68dca49622b3d8..5616cac6de37b4 100644
--- a/clang/test/SemaOpenACC/loop-construct.cpp
+++ b/clang/test/SemaOpenACC/loop-construct.cpp
@@ -28,7 +28,6 @@ struct SomeRAIterator {
int operator*();
void operator+=(int);
bool operator!=(SomeRAIterator&);
- // TODO
};
struct HasIteratorCollection {
More information about the cfe-commits
mailing list