Skip to content

Commit bbbaeb5

Browse files
erichkeaneerichkeane
committed
[OpenACC] Enable 'attach' clause for combined constructs
Once again, this clause has the same implementation for compute constructs as combined, so this adds the tests and enables it.
1 parent 35a5c71 commit bbbaeb5

6 files changed

+266
-23
lines changed

‎clang/lib/Sema/SemaOpenACC.cpp

Lines changed: 5 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -940,10 +940,11 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitCreateClause(
940940

941941
OpenACCClause *SemaOpenACCClauseVisitor::VisitAttachClause(
942942
SemaOpenACC::OpenACCParsedClause &Clause) {
943-
// Restrictions only properly implemented on 'compute' constructs, and
944-
// 'compute' constructs are the only construct that can do anything with
945-
// this yet, so skip/treat as unimplemented in this case.
946-
if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()))
943+
// Restrictions only properly implemented on 'compute'/'combined' constructs,
944+
// and 'compute'/'combined' constructs are the only construct that can do
945+
// anything with this yet, so skip/treat as unimplemented in this case.
946+
if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()) &&
947+
!isOpenACCCombinedDirectiveKind(Clause.getDirectiveKind()))
947948
return isNotImplemented();
948949

949950
// ActOnVar ensured that everything is a valid variable reference, but we

‎clang/test/AST/ast-print-openacc-combined-construct.cpp

Lines changed: 4 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -158,4 +158,8 @@ void foo() {
158158
#pragma acc parallel loop wait(devnum:i:queues:*iPtr, i)
159159
for(int i = 0;i<5;++i);
160160

161+
// CHECK: #pragma acc serial loop attach(iPtr, arrayPtr[0])
162+
#pragma acc serial loop attach(iPtr, arrayPtr[0])
163+
for(int i = 0;i<5;++i);
164+
161165
}

‎clang/test/SemaOpenACC/combined-construct-attach-ast.cpp

Lines changed: 66 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,66 @@
1+
// RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s
2+
3+
// Test this with PCH.
4+
// RUN: %clang_cc1 %s -fopenacc -emit-pch -o %t %s
5+
// RUN: %clang_cc1 %s -fopenacc -include-pch %t -ast-dump-all | FileCheck %s
6+
7+
#ifndef PCH_HELPER
8+
#define PCH_HELPER
9+
10+
void NormalUses(float *PointerParam) {
11+
// CHECK: FunctionDecl{{.*}}NormalUses
12+
// CHECK: ParmVarDecl
13+
// CHECK-NEXT: CompoundStmt
14+
15+
#pragma acc parallel loop attach(PointerParam) deviceptr(PointerParam)
16+
for (unsigned i = 0; i < 5; ++i);
17+
// CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
18+
// CHECK-NEXT: attach clause
19+
// CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}} 'PointerParam' 'float *'
20+
// CHECK-NEXT: deviceptr clause
21+
// CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}} 'PointerParam' 'float *'
22+
// CHECK-NEXT: ForStmt
23+
// CHECK: NullStmt
24+
}
25+
26+
template<typename T>
27+
void TemplUses(T *PointerParam) {
28+
// CHECK-NEXT: FunctionTemplateDecl
29+
// CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 0 T
30+
// CHECK-NEXT: FunctionDecl{{.*}} TemplUses 'void (T *)'
31+
// CHECK-NEXT: ParmVarDecl{{.*}} referenced PointerParam 'T *'
32+
// CHECK-NEXT: CompoundStmt
33+
34+
#pragma acc parallel loop attach(PointerParam) deviceptr(PointerParam)
35+
for (unsigned i = 0; i < 5; ++i);
36+
// CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
37+
// CHECK-NEXT: attach clause
38+
// CHECK-NEXT: DeclRefExpr{{.*}}'T *' lvalue ParmVar{{.*}} 'PointerParam' 'T *'
39+
// CHECK-NEXT: deviceptr clause
40+
// CHECK-NEXT: DeclRefExpr{{.*}}'T *' lvalue ParmVar{{.*}} 'PointerParam' 'T *'
41+
// CHECK-NEXT: ForStmt
42+
// CHECK: NullStmt
43+
44+
// Check the instantiated versions of the above.
45+
// CHECK-NEXT: FunctionDecl{{.*}} used TemplUses 'void (int *)' implicit_instantiation
46+
// CHECK-NEXT: TemplateArgument type 'int'
47+
// CHECK-NEXT: BuiltinType{{.*}} 'int'
48+
// CHECK-NEXT: ParmVarDecl{{.*}} used PointerParam 'int *'
49+
// CHECK-NEXT: CompoundStmt
50+
51+
//#pragma acc parallel loop attach(PointerParam) deviceptr(PointerParam)
52+
// CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop
53+
// CHECK-NEXT: attach clause
54+
// CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'PointerParam' 'int *'
55+
// CHECK-NEXT: deviceptr clause
56+
// CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'PointerParam' 'int *'
57+
// CHECK-NEXT: ForStmt
58+
// CHECK: NullStmt
59+
60+
}
61+
62+
void Inst() {
63+
int i;
64+
TemplUses(&i);
65+
}
66+
#endif

‎clang/test/SemaOpenACC/combined-construct-attach-clause.c

Lines changed: 65 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,65 @@
1+
// RUN: %clang_cc1 %s -fopenacc -verify
2+
3+
struct S {
4+
int IntMem;
5+
int *PtrMem;
6+
};
7+
8+
void uses() {
9+
int LocalInt;
10+
int *LocalPtr;
11+
int Array[5];
12+
int *PtrArray[5];
13+
struct S s;
14+
15+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int'}}
16+
#pragma acc parallel loop attach(LocalInt)
17+
for (unsigned i = 0; i < 5; ++i);
18+
19+
// expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}}
20+
#pragma acc parallel loop attach(&LocalInt)
21+
for (unsigned i = 0; i < 5; ++i);
22+
23+
#pragma acc serial loop attach(LocalPtr)
24+
for (unsigned i = 0; i < 5; ++i);
25+
26+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int[5]'}}
27+
#pragma acc kernels loop attach(Array)
28+
for (unsigned i = 0; i < 5; ++i);
29+
30+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int'}}
31+
#pragma acc parallel loop attach(Array[0])
32+
for (unsigned i = 0; i < 5; ++i);
33+
34+
// expected-error@+2{{OpenACC sub-array is not allowed here}}
35+
// expected-note@+1{{expected variable of pointer type}}
36+
#pragma acc parallel loop attach(Array[0:1])
37+
for (unsigned i = 0; i < 5; ++i);
38+
39+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int *[5]'}}
40+
#pragma acc parallel loop attach(PtrArray)
41+
for (unsigned i = 0; i < 5; ++i);
42+
43+
#pragma acc parallel loop attach(PtrArray[0])
44+
for (unsigned i = 0; i < 5; ++i);
45+
46+
// expected-error@+2{{OpenACC sub-array is not allowed here}}
47+
// expected-note@+1{{expected variable of pointer type}}
48+
#pragma acc parallel loop attach(PtrArray[0:1])
49+
for (unsigned i = 0; i < 5; ++i);
50+
51+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'struct S'}}
52+
#pragma acc parallel loop attach(s)
53+
for (unsigned i = 0; i < 5; ++i);
54+
55+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int'}}
56+
#pragma acc parallel loop attach(s.IntMem)
57+
for (unsigned i = 0; i < 5; ++i);
58+
59+
#pragma acc parallel loop attach(s.PtrMem)
60+
for (unsigned i = 0; i < 5; ++i);
61+
62+
// expected-error@+1{{OpenACC 'attach' clause is not valid on 'loop' directive}}
63+
#pragma acc loop attach(LocalInt)
64+
for(int i = 5; i < 10;++i);
65+
}

‎clang/test/SemaOpenACC/combined-construct-attach-clause.cpp

Lines changed: 120 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,120 @@
1+
// RUN: %clang_cc1 %s -fopenacc -verify
2+
3+
struct S {
4+
int IntMem;
5+
int *PtrMem;
6+
operator int*();
7+
};
8+
9+
void uses() {
10+
int LocalInt;
11+
int *LocalPtr;
12+
int Array[5];
13+
int *PtrArray[5];
14+
struct S s;
15+
16+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int'}}
17+
#pragma acc parallel loop attach(LocalInt)
18+
for (unsigned i = 0; i < 5; ++i);
19+
20+
#pragma acc parallel loop attach(LocalPtr)
21+
for (unsigned i = 0; i < 5; ++i);
22+
23+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int[5]'}}
24+
#pragma acc parallel loop attach(Array)
25+
for (unsigned i = 0; i < 5; ++i);
26+
27+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int'}}
28+
#pragma acc parallel loop attach(Array[0])
29+
for (unsigned i = 0; i < 5; ++i);
30+
31+
// expected-error@+2{{OpenACC sub-array is not allowed here}}
32+
// expected-note@+1{{expected variable of pointer type}}
33+
#pragma acc parallel loop attach(Array[0:1])
34+
for (unsigned i = 0; i < 5; ++i);
35+
36+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int *[5]'}}
37+
#pragma acc parallel loop attach(PtrArray)
38+
for (unsigned i = 0; i < 5; ++i);
39+
40+
#pragma acc parallel loop attach(PtrArray[0])
41+
for (unsigned i = 0; i < 5; ++i);
42+
43+
// expected-error@+2{{OpenACC sub-array is not allowed here}}
44+
// expected-note@+1{{expected variable of pointer type}}
45+
#pragma acc parallel loop attach(PtrArray[0:1])
46+
for (unsigned i = 0; i < 5; ++i);
47+
48+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'struct S'}}
49+
#pragma acc parallel loop attach(s)
50+
for (unsigned i = 0; i < 5; ++i);
51+
52+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int'}}
53+
#pragma acc parallel loop attach(s.IntMem)
54+
for (unsigned i = 0; i < 5; ++i);
55+
56+
#pragma acc parallel loop attach(s.PtrMem)
57+
for (unsigned i = 0; i < 5; ++i);
58+
}
59+
60+
template<typename T, typename TPtr, typename TStruct, auto &R1>
61+
void Templ() {
62+
T SomeInt;
63+
TPtr SomePtr;
64+
T SomeIntArray[5];
65+
TPtr SomeIntPtrArray[5];
66+
TStruct SomeStruct;
67+
68+
// expected-error@+2{{expected pointer in 'attach' clause, type is 'int'}}
69+
// expected-note@#INST{{in instantiation of function template specialization}}
70+
#pragma acc parallel loop attach(SomeInt)
71+
for (unsigned i = 0; i < 5; ++i);
72+
73+
#pragma acc parallel loop attach(SomePtr)
74+
for (unsigned i = 0; i < 5; ++i);
75+
76+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int[5]'}}
77+
#pragma acc parallel loop attach(SomeIntArray)
78+
for (unsigned i = 0; i < 5; ++i);
79+
80+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int'}}
81+
#pragma acc parallel loop attach(SomeIntArray[0])
82+
for (unsigned i = 0; i < 5; ++i);
83+
84+
// expected-error@+2{{OpenACC sub-array is not allowed here}}
85+
// expected-note@+1{{expected variable of pointer type}}
86+
#pragma acc parallel loop attach(SomeIntArray[0:1])
87+
for (unsigned i = 0; i < 5; ++i);
88+
89+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int *[5]'}}
90+
#pragma acc parallel loop attach(SomeIntPtrArray)
91+
for (unsigned i = 0; i < 5; ++i);
92+
93+
#pragma acc parallel loop attach(SomeIntPtrArray[0])
94+
for (unsigned i = 0; i < 5; ++i);
95+
96+
// expected-error@+2{{OpenACC sub-array is not allowed here}}
97+
// expected-note@+1{{expected variable of pointer type}}
98+
#pragma acc parallel loop attach(SomeIntPtrArray[0:1])
99+
for (unsigned i = 0; i < 5; ++i);
100+
101+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'S'}}
102+
#pragma acc parallel loop attach(SomeStruct)
103+
for (unsigned i = 0; i < 5; ++i);
104+
105+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int'}}
106+
#pragma acc parallel loop attach(SomeStruct.IntMem)
107+
for (unsigned i = 0; i < 5; ++i);
108+
109+
#pragma acc parallel loop attach(SomeStruct.PtrMem)
110+
for (unsigned i = 0; i < 5; ++i);
111+
112+
// expected-error@+1{{expected pointer in 'attach' clause, type is 'int'}}
113+
#pragma acc parallel loop attach(R1)
114+
for (unsigned i = 0; i < 5; ++i);
115+
}
116+
117+
void inst() {
118+
static constexpr int CEVar = 1;
119+
Templ<int, int*, S, CEVar>(); // #INST
120+
}

‎clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c

Lines changed: 6 additions & 19 deletions
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,3 @@
1-
21
// RUN: %clang_cc1 %s -fopenacc -verify
32

43
// TODO: OpenACC: A number of the 'not yet implemented' diagnostics interfere
@@ -79,9 +78,7 @@ void uses() {
7978
// expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}}
8079
#pragma acc parallel loop auto use_device(Var)
8180
for(unsigned i = 0; i < 5; ++i);
82-
// TODOexpected-error@+1{{OpenACC 'attach' clause is not valid on 'parallel loop' directive}}
83-
// expected-warning@+1{{OpenACC clause 'attach' not yet implemented}}
84-
#pragma acc parallel loop auto attach(Var)
81+
#pragma acc parallel loop auto attach(VarPtr)
8582
for(unsigned i = 0; i < 5; ++i);
8683
// expected-warning@+1{{OpenACC clause 'delete' not yet implemented}}
8784
#pragma acc parallel loop auto delete(Var)
@@ -236,9 +233,7 @@ void uses() {
236233
// expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}}
237234
#pragma acc parallel loop use_device(Var) auto
238235
for(unsigned i = 0; i < 5; ++i);
239-
// TODOexpected-error@+1{{OpenACC 'attach' clause is not valid on 'parallel loop' directive}}
240-
// expected-warning@+1{{OpenACC clause 'attach' not yet implemented}}
241-
#pragma acc parallel loop attach(Var) auto
236+
#pragma acc parallel loop attach(VarPtr) auto
242237
for(unsigned i = 0; i < 5; ++i);
243238
// expected-warning@+1{{OpenACC clause 'delete' not yet implemented}}
244239
#pragma acc parallel loop delete(Var) auto
@@ -394,9 +389,7 @@ void uses() {
394389
// expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}}
395390
#pragma acc parallel loop independent use_device(Var)
396391
for(unsigned i = 0; i < 5; ++i);
397-
// TODOexpected-error@+1{{OpenACC 'attach' clause is not valid on 'parallel loop' directive}}
398-
// expected-warning@+1{{OpenACC clause 'attach' not yet implemented}}
399-
#pragma acc parallel loop independent attach(Var)
392+
#pragma acc parallel loop independent attach(VarPtr)
400393
for(unsigned i = 0; i < 5; ++i);
401394
// expected-warning@+1{{OpenACC clause 'delete' not yet implemented}}
402395
#pragma acc parallel loop independent delete(Var)
@@ -551,9 +544,7 @@ void uses() {
551544
// expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}}
552545
#pragma acc parallel loop use_device(Var) independent
553546
for(unsigned i = 0; i < 5; ++i);
554-
// TODOexpected-error@+1{{OpenACC 'attach' clause is not valid on 'parallel loop' directive}}
555-
// expected-warning@+1{{OpenACC clause 'attach' not yet implemented}}
556-
#pragma acc parallel loop attach(Var) independent
547+
#pragma acc parallel loop attach(VarPtr) independent
557548
for(unsigned i = 0; i < 5; ++i);
558549
// expected-warning@+1{{OpenACC clause 'delete' not yet implemented}}
559550
#pragma acc parallel loop delete(Var) independent
@@ -715,9 +706,7 @@ void uses() {
715706
// expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}}
716707
#pragma acc parallel loop seq use_device(Var)
717708
for(unsigned i = 0; i < 5; ++i);
718-
// TODOexpected-error@+1{{OpenACC 'attach' clause is not valid on 'parallel loop' directive}}
719-
// expected-warning@+1{{OpenACC clause 'attach' not yet implemented}}
720-
#pragma acc parallel loop seq attach(Var)
709+
#pragma acc parallel loop seq attach(VarPtr)
721710
for(unsigned i = 0; i < 5; ++i);
722711
// expected-warning@+1{{OpenACC clause 'delete' not yet implemented}}
723712
#pragma acc parallel loop seq delete(Var)
@@ -878,9 +867,7 @@ void uses() {
878867
// expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}}
879868
#pragma acc parallel loop use_device(Var) seq
880869
for(unsigned i = 0; i < 5; ++i);
881-
// TODOexpected-error@+1{{OpenACC 'attach' clause is not valid on 'parallel loop' directive}}
882-
// expected-warning@+1{{OpenACC clause 'attach' not yet implemented}}
883-
#pragma acc parallel loop attach(Var) seq
870+
#pragma acc parallel loop attach(VarPtr) seq
884871
for(unsigned i = 0; i < 5; ++i);
885872
// expected-warning@+1{{OpenACC clause 'delete' not yet implemented}}
886873
#pragma acc parallel loop delete(Var) seq

0 commit comments

Comments
 (0)