mirrors/llvm-project
0
0
Fork 0
mirror of https://github.com/llvm/llvm-project.git synced 2025-04-26 11:16:09 +00:00
Code Issues Projects Releases Wiki Activity

[OpenACC] Remove 'loop' link to parent construct

Browse Source 
After implementing 'loop', we determined that the link to its parent
only ever uses the type, not the construct itself. This patch removes
it, as it is both a waste and causes problems with serialization.
This commit is contained in:
erichkeane 2024-11-08 09:57:19 -08:00
parent 17f3e00911
commit e4d57d6a72
13 changed files with 132 additions and 195 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

37
clang/include/clang/AST/StmtOpenACC.h
View File

@ -114,7 +114,6 @@ public:
}
};
class OpenACCLoopConstruct;
/// This class represents a compute construct, representing a 'Kind' of
/// `parallel', 'serial', or 'kernel'. These constructs are associated with a
/// 'structured block', defined as:
@ -183,8 +182,7 @@ public:
static OpenACCComputeConstruct *
Create(const ASTContext &C, OpenACCDirectiveKind K, SourceLocation BeginLoc,
SourceLocation DirectiveLoc, SourceLocation EndLoc,
ArrayRef<const OpenACCClause *> Clauses, Stmt *StructuredBlock,
ArrayRef<OpenACCLoopConstruct *> AssociatedLoopConstructs);
ArrayRef<const OpenACCClause *> Clauses, Stmt *StructuredBlock);
Stmt *getStructuredBlock() { return getAssociatedStmt(); }
const Stmt *getStructuredBlock() const {
@ -198,12 +196,10 @@ class OpenACCLoopConstruct final
: public OpenACCAssociatedStmtConstruct,
public llvm::TrailingObjects<OpenACCLoopConstruct,
const OpenACCClause *> {
// The compute construct this loop is associated with, or nullptr if this is
// an orphaned loop construct, or if it hasn't been set yet. Because we
// construct the directives at the end of their statement, the 'parent'
// construct is not yet available at the time of construction, so this needs
// to be set 'later'.
const OpenACCComputeConstruct *ParentComputeConstruct = nullptr;
// The compute/combined construct kind this loop is associated with, or
// invalid if this is an orphaned loop construct.
OpenACCDirectiveKind ParentComputeConstructKind =
OpenACCDirectiveKind::Invalid;
friend class ASTStmtWriter;
friend class ASTStmtReader;
@ -212,15 +208,9 @@ class OpenACCLoopConstruct final
OpenACCLoopConstruct(unsigned NumClauses);
OpenACCLoopConstruct(SourceLocation Start, SourceLocation DirLoc,
SourceLocation End,
OpenACCLoopConstruct(OpenACCDirectiveKind ParentKind, SourceLocation Start,
SourceLocation DirLoc, SourceLocation End,
ArrayRef<const OpenACCClause *> Clauses, Stmt *Loop);
void setLoop(Stmt *Loop);
void setParentComputeConstruct(OpenACCComputeConstruct *CC) {
assert(!ParentComputeConstruct && "Parent already set?");
ParentComputeConstruct = CC;
}
public:
static bool classof(const Stmt *T) {
@ -231,9 +221,9 @@ public:
unsigned NumClauses);
static OpenACCLoopConstruct *
Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation DirLoc,
SourceLocation EndLoc, ArrayRef<const OpenACCClause *> Clauses,
Stmt *Loop);
Create(const ASTContext &C, OpenACCDirectiveKind ParentKind,
SourceLocation BeginLoc, SourceLocation DirLoc, SourceLocation EndLoc,
ArrayRef<const OpenACCClause *> Clauses, Stmt *Loop);
Stmt *getLoop() { return getAssociatedStmt(); }
const Stmt *getLoop() const {
@ -246,10 +236,11 @@ public:
/// loop construct is the nearest compute construct that lexically contains
/// the loop construct.
bool isOrphanedLoopConstruct() const {
return ParentComputeConstruct == nullptr;
return ParentComputeConstructKind == OpenACCDirectiveKind::Invalid;
}
const OpenACCComputeConstruct *getParentComputeConstruct() const {
return ParentComputeConstruct;
OpenACCDirectiveKind getParentComputeConstructKind() const {
return ParentComputeConstructKind;
}
};
} // namespace clang

6
clang/include/clang/Sema/SemaOpenACC.h
View File

@ -34,11 +34,6 @@ class OpenACCClause;
class SemaOpenACC : public SemaBase {
private:
/// A collection of loop constructs in the compute construct scope that
/// haven't had their 'parent' compute construct set yet. Entires will only be
/// made to this list in the case where we know the loop isn't an orphan.
llvm::SmallVector<OpenACCLoopConstruct *> ParentlessLoopConstructs;
struct ComputeConstructInfo {
/// Which type of compute construct we are inside of, which we can use to
/// determine whether we should add loops to the above collection. We can
@ -768,7 +763,6 @@ public:
SourceLocation OldLoopWorkerClauseLoc;
SourceLocation OldLoopVectorClauseLoc;
SourceLocation OldLoopWithoutSeqLoc;
llvm::SmallVector<OpenACCLoopConstruct *> ParentlessLoopConstructs;
llvm::SmallVector<OpenACCReductionClause *> ActiveReductionClauses;
LoopInConstructRAII LoopRAII;

56
clang/lib/AST/StmtOpenACC.cpp
View File

@ -28,44 +28,15 @@ OpenACCComputeConstruct::CreateEmpty(const ASTContext &C, unsigned NumClauses) {
OpenACCComputeConstruct *OpenACCComputeConstruct::Create(
const ASTContext &C, OpenACCDirectiveKind K, SourceLocation BeginLoc,
SourceLocation DirLoc, SourceLocation EndLoc,
ArrayRef<const OpenACCClause *> Clauses, Stmt *StructuredBlock,
ArrayRef<OpenACCLoopConstruct *> AssociatedLoopConstructs) {
ArrayRef<const OpenACCClause *> Clauses, Stmt *StructuredBlock) {
void *Mem = C.Allocate(
OpenACCComputeConstruct::totalSizeToAlloc<const OpenACCClause *>(
Clauses.size()));
auto *Inst = new (Mem) OpenACCComputeConstruct(K, BeginLoc, DirLoc, EndLoc,
Clauses, StructuredBlock);
llvm::for_each(AssociatedLoopConstructs, [&](OpenACCLoopConstruct *C) {
C->setParentComputeConstruct(Inst);
});
return Inst;
}
void OpenACCComputeConstruct::findAndSetChildLoops() {
struct LoopConstructFinder : RecursiveASTVisitor<LoopConstructFinder> {
OpenACCComputeConstruct *Construct = nullptr;
LoopConstructFinder(OpenACCComputeConstruct *Construct)
: Construct(Construct) {}
bool TraverseOpenACCComputeConstruct(OpenACCComputeConstruct *C) {
// Stop searching if we find a compute construct.
return true;
}
bool TraverseOpenACCLoopConstruct(OpenACCLoopConstruct *C) {
// Stop searching if we find a loop construct, after taking ownership of
// it.
C->setParentComputeConstruct(Construct);
return true;
}
};
LoopConstructFinder f(this);
f.TraverseStmt(getAssociatedStmt());
}
OpenACCLoopConstruct::OpenACCLoopConstruct(unsigned NumClauses)
: OpenACCAssociatedStmtConstruct(
OpenACCLoopConstructClass, OpenACCDirectiveKind::Loop,
@ -79,11 +50,13 @@ OpenACCLoopConstruct::OpenACCLoopConstruct(unsigned NumClauses)
}
OpenACCLoopConstruct::OpenACCLoopConstruct(
SourceLocation Start, SourceLocation DirLoc, SourceLocation End,
OpenACCDirectiveKind ParentKind, SourceLocation Start,
SourceLocation DirLoc, SourceLocation End,
ArrayRef<const OpenACCClause *> Clauses, Stmt *Loop)
: OpenACCAssociatedStmtConstruct(OpenACCLoopConstructClass,
OpenACCDirectiveKind::Loop, Start, DirLoc,
End, Loop) {
End, Loop),
ParentComputeConstructKind(ParentKind) {
// accept 'nullptr' for the loop. This is diagnosed somewhere, but this gives
// us some level of AST fidelity in the error case.
assert((Loop == nullptr || isa<ForStmt, CXXForRangeStmt>(Loop)) &&
@ -96,12 +69,6 @@ OpenACCLoopConstruct::OpenACCLoopConstruct(
Clauses.size()));
}
void OpenACCLoopConstruct::setLoop(Stmt *Loop) {
assert((isa<ForStmt, CXXForRangeStmt>(Loop)) &&
"Associated Loop not a for loop?");
setAssociatedStmt(Loop);
}
OpenACCLoopConstruct *OpenACCLoopConstruct::CreateEmpty(const ASTContext &C,
unsigned NumClauses) {
void *Mem =
@ -111,15 +78,14 @@ OpenACCLoopConstruct *OpenACCLoopConstruct::CreateEmpty(const ASTContext &C,
return Inst;
}
OpenACCLoopConstruct *
OpenACCLoopConstruct::Create(const ASTContext &C, SourceLocation BeginLoc,
SourceLocation DirLoc, SourceLocation EndLoc,
ArrayRef<const OpenACCClause *> Clauses,
Stmt *Loop) {
OpenACCLoopConstruct *OpenACCLoopConstruct::Create(
const ASTContext &C, OpenACCDirectiveKind ParentKind,
SourceLocation BeginLoc, SourceLocation DirLoc, SourceLocation EndLoc,
ArrayRef<const OpenACCClause *> Clauses, Stmt *Loop) {
void *Mem =
C.Allocate(OpenACCLoopConstruct::totalSizeToAlloc<const OpenACCClause *>(
Clauses.size()));
auto *Inst =
new (Mem) OpenACCLoopConstruct(BeginLoc, DirLoc, EndLoc, Clauses, Loop);
auto *Inst = new (Mem)
OpenACCLoopConstruct(ParentKind, BeginLoc, DirLoc, EndLoc, Clauses, Loop);
return Inst;
}

2
clang/lib/AST/TextNodeDumper.cpp
View File

@ -2928,7 +2928,7 @@ void TextNodeDumper::VisitOpenACCLoopConstruct(const OpenACCLoopConstruct *S) {
if (S->isOrphanedLoopConstruct())
OS << " <orphan>";
else
OS << " parent: " << S->getParentComputeConstruct();
OS << " parent: " << S->getParentComputeConstructKind();
}
void TextNodeDumper::VisitEmbedExpr(const EmbedExpr *S) {

29
clang/lib/Sema/SemaOpenACC.cpp
View File

@ -1538,7 +1538,6 @@ SemaOpenACC::AssociatedStmtRAII::AssociatedStmtRAII(
CollectActiveReductionClauses(S.ActiveReductionClauses, Clauses);
SemaRef.ActiveComputeConstructInfo.Kind = DirKind;
SemaRef.ActiveComputeConstructInfo.Clauses = Clauses;
SemaRef.ParentlessLoopConstructs.swap(ParentlessLoopConstructs);
// OpenACC 3.3 2.9.2: When the parent compute construct is a kernels
// construct, the gang clause behaves as follows. ... The region of a loop
@ -1668,9 +1667,8 @@ SemaOpenACC::AssociatedStmtRAII::~AssociatedStmtRAII() {
if (DirKind == OpenACCDirectiveKind::Parallel ||
DirKind == OpenACCDirectiveKind::Serial ||
DirKind == OpenACCDirectiveKind::Kernels) {
assert(SemaRef.ParentlessLoopConstructs.empty() &&
"Didn't consume loop construct list?");
SemaRef.ParentlessLoopConstructs.swap(ParentlessLoopConstructs);
// Nothing really to do here, the restorations above should be enough for
// now.
} else if (DirKind == OpenACCDirectiveKind::Loop) {
// Nothing really to do here, the LoopInConstruct should handle restorations
// correctly.
@ -3171,27 +3169,14 @@ StmtResult SemaOpenACC::ActOnEndStmtDirective(OpenACCDirectiveKind K,
case OpenACCDirectiveKind::Parallel:
case OpenACCDirectiveKind::Serial:
case OpenACCDirectiveKind::Kernels: {
auto *ComputeConstruct = OpenACCComputeConstruct::Create(
return OpenACCComputeConstruct::Create(
getASTContext(), K, StartLoc, DirLoc, EndLoc, Clauses,
AssocStmt.isUsable() ? AssocStmt.get() : nullptr,
ParentlessLoopConstructs);
ParentlessLoopConstructs.clear();
return ComputeConstruct;
AssocStmt.isUsable() ? AssocStmt.get() : nullptr);
}
case OpenACCDirectiveKind::Loop: {
auto *LoopConstruct = OpenACCLoopConstruct::Create(
getASTContext(), StartLoc, DirLoc, EndLoc, Clauses,
AssocStmt.isUsable() ? AssocStmt.get() : nullptr);
// If we are in the scope of a compute construct, add this to the list of
// loop constructs that need assigning to the next closing compute
// construct.
if (isInComputeConstruct())
ParentlessLoopConstructs.push_back(LoopConstruct);
return LoopConstruct;
return OpenACCLoopConstruct::Create(
getASTContext(), ActiveComputeConstructInfo.Kind, StartLoc, DirLoc,
EndLoc, Clauses, AssocStmt.isUsable() ? AssocStmt.get() : nullptr);
}
}
llvm_unreachable("Unhandled case in directive handling?");

2
clang/lib/Serialization/ASTReaderStmt.cpp
View File

@ -2836,12 +2836,12 @@ void ASTStmtReader::VisitOpenACCAssociatedStmtConstruct(
void ASTStmtReader::VisitOpenACCComputeConstruct(OpenACCComputeConstruct *S) {
VisitStmt(S);
VisitOpenACCAssociatedStmtConstruct(S);
S->findAndSetChildLoops();
}
void ASTStmtReader::VisitOpenACCLoopConstruct(OpenACCLoopConstruct *S) {
VisitStmt(S);
VisitOpenACCAssociatedStmtConstruct(S);
S->ParentComputeConstructKind = Record.readEnum<OpenACCDirectiveKind>();
}
//===----------------------------------------------------------------------===//

1
clang/lib/Serialization/ASTWriterStmt.cpp
View File

@ -2915,6 +2915,7 @@ void ASTStmtWriter::VisitOpenACCComputeConstruct(OpenACCComputeConstruct *S) {
void ASTStmtWriter::VisitOpenACCLoopConstruct(OpenACCLoopConstruct *S) {
VisitStmt(S);
VisitOpenACCAssociatedStmtConstruct(S);
Record.writeEnum(S->getParentComputeConstructKind());
Code = serialization::STMT_OPENACC_LOOP_CONSTRUCT;
}

6
clang/test/SemaOpenACC/compute-construct-default-clause.c
View File

@ -23,18 +23,18 @@ void SingleOnly() {
// expected-warning@+2{{OpenACC clause 'default' not yet implemented}}
// expected-warning@+1{{OpenACC clause 'copy' not yet implemented}}
#pragma acc parallel loop self default(present) private(i) default(none) copy(i)
while(0);
for(int i = 0; i < 5; ++i);
// expected-warning@+3{{OpenACC clause 'self' not yet implemented, clause ignored}}
// expected-warning@+2{{OpenACC construct 'serial loop' not yet implemented}}
// expected-error@+1{{expected '('}}
#pragma acc serial loop self default private(i) default(none) if(i)
while(0);
for(int i = 0; i < 5; ++i);
// expected-warning@+2{{OpenACC construct 'kernels loop' not yet implemented}}
// expected-warning@+1{{OpenACC clause 'default' not yet implemented}}
#pragma acc kernels loop default(none)
while(0);
for(int i = 0; i < 5; ++i);
// expected-warning@+2{{OpenACC construct 'data' not yet implemented}}
// expected-warning@+1{{OpenACC clause 'default' not yet implemented}}

6
clang/test/SemaOpenACC/compute-construct-if-clause.c
View File

@ -50,15 +50,15 @@ void BoolExpr(int *I, float *F) {
// expected-warning@+2{{OpenACC construct 'parallel loop' not yet implemented}}
// expected-warning@+1{{OpenACC clause 'if' not yet implemented}}
#pragma acc parallel loop if (*I < *F)
while(0);
for(int i = 0; i < 5; ++i);
// expected-warning@+2{{OpenACC construct 'serial loop' not yet implemented}}
// expected-warning@+1{{OpenACC clause 'if' not yet implemented}}
#pragma acc serial loop if (*I < *F)
while(0);
for(int i = 0; i < 5; ++i);
// expected-warning@+2{{OpenACC construct 'kernels loop' not yet implemented}}
// expected-warning@+1{{OpenACC clause 'if' not yet implemented}}
#pragma acc kernels loop if (*I < *F)
while(0);
for(int i = 0; i < 5; ++i);
// expected-error@+1{{OpenACC 'if' clause is not valid on 'loop' directive}}
#pragma acc loop if(I)

22
clang/test/SemaOpenACC/loop-ast.cpp
View File

@ -42,12 +42,12 @@ void NormalFunc() {
// CHECK-NEXT: CompoundStmt
{
#pragma acc parallel
// CHECK-NEXT: OpenACCComputeConstruct [[PAR_ADDR:[0-9a-fx]+]] {{.*}}parallel
// CHECK-NEXT: OpenACCComputeConstruct {{.*}}parallel
// CHECK-NEXT: CompoundStmt
{
#pragma acc loop
for(int i = 0; i < 5;++i);
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR]]
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
// CHECK-NEXT: VarDecl {{.*}} used i 'int'
@ -91,16 +91,16 @@ void TemplFunc() {
}
#pragma acc parallel
#pragma acc serial
{
// CHECK-NEXT: OpenACCComputeConstruct {{.*}}parallel
// CHECK-NEXT: OpenACCComputeConstruct {{.*}}serial
// CHECK-NEXT: CompoundStmt
#pragma acc parallel
{
// CHECK-NEXT: OpenACCComputeConstruct [[PAR_ADDR_UNINST:[0-9a-fx]+]] {{.*}}parallel
// CHECK-NEXT: OpenACCComputeConstruct {{.*}}parallel
// CHECK-NEXT: CompoundStmt
#pragma acc loop
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_UNINST]]
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
// CHECK-NEXT: VarDecl {{.*}} i 'int'
@ -116,7 +116,7 @@ void TemplFunc() {
for(int i = 0; i < 5;++i);
#pragma acc loop
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_UNINST]]
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
// CHECK-NEXT: VarDecl {{.*}} i 'int'
@ -166,13 +166,13 @@ void TemplFunc() {
// CHECK-NEXT: DeclStmt
// CHECK-NEXT: VarDecl{{.*}} I 'typename S::type':'int'
// CHECK-NEXT: OpenACCComputeConstruct {{.*}}parallel
// CHECK-NEXT: OpenACCComputeConstruct {{.*}}serial
// CHECK-NEXT: CompoundStmt
//
// CHECK-NEXT: OpenACCComputeConstruct [[PAR_ADDR_INST:[0-9a-fx]+]] {{.*}}parallel
// CHECK-NEXT: OpenACCComputeConstruct {{.*}}parallel
// CHECK-NEXT: CompoundStmt
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_INST]]
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
// CHECK-NEXT: VarDecl {{.*}} i 'int'
@ -186,7 +186,7 @@ void TemplFunc() {
// CHECK-NEXT: DeclRefExpr{{.*}} 'i' 'int'
// CHECK-NEXT: NullStmt
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_INST]]
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
// CHECK-NEXT: VarDecl {{.*}} i 'int'

52
clang/test/SemaOpenACC/loop-construct-gang-ast.cpp
View File

@ -53,8 +53,8 @@ void NormalUses() {
#pragma acc loop gang(static:Val)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: gang clause num
// CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
// CHECK-NEXT: gang clause static
@ -76,8 +76,8 @@ void NormalUses() {
#pragma acc loop gang(num:1) gang(static:Val)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: gang clause dim static
// CHECK-NEXT: ConstantExpr{{.*}} 'int'
// CHECK-NEXT: value: Int 1
@ -100,8 +100,8 @@ void NormalUses() {
#pragma acc loop gang(dim:1, static:Val)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: gang clause static
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}} 'Val' 'int'
@ -121,8 +121,8 @@ void NormalUses() {
#pragma acc loop gang(static:Val)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}}serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: gang clause static
// CHECK-NEXT: OpenACCAsteriskSizeExpr
// CHECK-NEXT: ForStmt
@ -141,8 +141,8 @@ void NormalUses() {
#pragma acc loop gang(static:*)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: gang clause
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -224,8 +224,8 @@ void TemplateUses(T Val) {
#pragma acc loop gang(static:*)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: gang clause dim
// CHECK-NEXT: DeclRefExpr{{.*}}'unsigned int' NonTypeTemplateParm{{.*}} 'One' 'unsigned int'
// CHECK-NEXT: gang clause static
@ -246,8 +246,8 @@ void TemplateUses(T Val) {
#pragma acc loop gang(dim:One) gang(static:Val)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: gang clause dim static
// CHECK-NEXT: DeclRefExpr{{.*}}'unsigned int' NonTypeTemplateParm{{.*}} 'One' 'unsigned int'
// CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 'Val' 'T'
@ -267,8 +267,8 @@ void TemplateUses(T Val) {
#pragma acc loop gang(dim:One, static:Val)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: gang clause static
// CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 'Val' 'T'
// CHECK-NEXT: ForStmt
@ -287,8 +287,8 @@ void TemplateUses(T Val) {
#pragma acc loop gang(static:Val)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: gang clause
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -367,8 +367,8 @@ void TemplateUses(T Val) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: gang clause dim
// CHECK-NEXT: ConstantExpr{{.*}} 'unsigned int'
// CHECK-NEXT: value: Int 1
@ -391,8 +391,8 @@ void TemplateUses(T Val) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: gang clause dim static
// CHECK-NEXT: ConstantExpr{{.*}} 'unsigned int'
// CHECK-NEXT: value: Int 1
@ -414,8 +414,8 @@ void TemplateUses(T Val) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: gang clause static
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 'Val' 'int'
@ -432,8 +432,8 @@ void TemplateUses(T Val) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: gang clause
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt

52
clang/test/SemaOpenACC/loop-construct-vector-ast.cpp
View File

@ -70,8 +70,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
#pragma acc loop vector(length:CTI)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: vector clause
// CHECK-NEXT: DeclRefExpr{{.*}}'Int' lvalue ParmVar{{.*}}'IsI' 'Int'
// CHECK-NEXT: ForStmt
@ -90,8 +90,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
#pragma acc loop vector(length:IsI)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: vector clause
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -109,8 +109,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
#pragma acc loop vector
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: vector clause
// CHECK-NEXT: DeclRefExpr{{.*}}'Int' lvalue ParmVar{{.*}}'IsI' 'Int'
// CHECK-NEXT: ForStmt
@ -194,8 +194,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: vector clause
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}}'IsI' 'int'
@ -212,8 +212,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: vector clause
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -228,8 +228,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: vector clause
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}}'IsI' 'int'
@ -330,8 +330,8 @@ void uses() {
#pragma acc loop vector(length:C)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: vector clause
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -349,8 +349,8 @@ void uses() {
#pragma acc loop vector
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: vector clause
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <UserDefinedConversion>
// CHECK-NEXT: CXXMemberCallExpr{{.*}}'int'
@ -372,8 +372,8 @@ void uses() {
#pragma acc loop vector(C)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: vector clause
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var
@ -393,8 +393,8 @@ void uses() {
#pragma acc loop vector(length:i)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: vector clause
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -412,8 +412,8 @@ void uses() {
#pragma acc loop vector
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: vector clause
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <UserDefinedConversion>
// CHECK-NEXT: CXXMemberCallExpr{{.*}}'int'
@ -435,8 +435,8 @@ void uses() {
#pragma acc loop vector(C)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: vector clause
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var
@ -456,8 +456,8 @@ void uses() {
#pragma acc loop vector(length:i)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: vector clause
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt

56
clang/test/SemaOpenACC/loop-construct-worker-ast.cpp
View File

@ -34,8 +34,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
#pragma acc loop worker
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -53,8 +53,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
#pragma acc loop worker
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -72,8 +72,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
#pragma acc loop worker
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: DeclRefExpr{{.*}} 'ConvertsToInt' lvalue ParmVar
// CHECK-NEXT: ForStmt
@ -92,8 +92,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
#pragma acc loop worker(CTI)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: DeclRefExpr{{.*}} 'Int' lvalue ParmVar
// CHECK-NEXT: ForStmt
@ -112,8 +112,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
#pragma acc loop worker(num:IsI)
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}}'I' 'unsigned int'
// CHECK-NEXT: ForStmt
@ -159,8 +159,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -175,8 +175,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -191,8 +191,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <UserDefinedConversion>
// CHECK-NEXT: CXXMemberCallExpr{{.*}} 'int'
@ -211,8 +211,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'IsI' 'int'
@ -229,8 +229,8 @@ void TemplUses(ConvertsToInt CTI, Int IsI) {
// CHECK-NEXT: DeclRefExpr{{.*}}'i' 'int'
// CHECK-NEXT: NullStmt
//
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}}'unsigned int'
// CHECK-NEXT: NonTypeTemplateParmDecl{{.*}}'unsigned int' depth 0 index 0 I
@ -277,8 +277,8 @@ void uses() {
#pragma acc loop worker
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} parallel
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: parallel
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -296,8 +296,8 @@ void uses() {
#pragma acc loop worker
for(int i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} serial
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: serial
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ForStmt
// CHECK-NEXT: DeclStmt
@ -320,8 +320,8 @@ void uses() {
// CHECK-NEXT: VarDecl
// CHECK-NEXT: CXXConstructExpr
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <UserDefinedConversion>
// CHECK-NEXT: CXXMemberCallExpr{{.*}} 'int'
@ -347,8 +347,8 @@ void uses() {
// CHECK-NEXT: DeclStmt
// CHECK-NEXT: VarDecl
// CHECK-NEXT: OpenACCComputeConstruct 0x[[COMPUTE_ADDR:[0-9a-f]+]]{{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: 0x[[COMPUTE_ADDR]]
// CHECK-NEXT: OpenACCComputeConstruct {{.*}} kernels
// CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: kernels
// CHECK-NEXT: worker clause{{.*}}
// CHECK-NEXT: ImplicitCastExpr{{.*}}'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var