Skip to content

Commit f9286b4

Browse files
jdoerfertjdoerfert
committed
[OpenMP] Attribute target diagnostics properly
Type errors in function declarations were not (always) diagnosed prior to this patch. Furthermore, certain remarks did not get associated properly which caused them to be emitted multiple times. Reviewed By: JonChesterfield Differential Revision: https://reviews.llvm.org/D95912
1 parent 3b2f19d commit f9286b4

File tree

6 files changed

+87
-48
lines changed

6 files changed

+87
-48
lines changed

‎clang/include/clang/Sema/Sema.h

Lines changed: 9 additions & 7 deletions
Original file line numberDiff line numberDiff line change
@@ -11951,8 +11951,8 @@ class Sema final {
1195111951
/// if (diagIfOpenMPDeviceCode(Loc, diag::err_vla_unsupported))
1195211952
/// return ExprError();
1195311953
/// // Otherwise, continue parsing as normal.
11954-
SemaDiagnosticBuilder diagIfOpenMPDeviceCode(SourceLocation Loc,
11955-
unsigned DiagID);
11954+
SemaDiagnosticBuilder
11955+
diagIfOpenMPDeviceCode(SourceLocation Loc, unsigned DiagID, FunctionDecl *FD);
1195611956

1195711957
/// Creates a SemaDiagnosticBuilder that emits the diagnostic if the current
1195811958
/// context is "used as host code".
@@ -11968,17 +11968,19 @@ class Sema final {
1196811968
/// return ExprError();
1196911969
/// // Otherwise, continue parsing as normal.
1197011970
SemaDiagnosticBuilder diagIfOpenMPHostCode(SourceLocation Loc,
11971-
unsigned DiagID);
11971+
unsigned DiagID, FunctionDecl *FD);
1197211972

11973-
SemaDiagnosticBuilder targetDiag(SourceLocation Loc, unsigned DiagID);
11973+
SemaDiagnosticBuilder targetDiag(SourceLocation Loc, unsigned DiagID,
11974+
FunctionDecl *FD = nullptr);
1197411975
SemaDiagnosticBuilder targetDiag(SourceLocation Loc,
11975-
const PartialDiagnostic &PD) {
11976-
return targetDiag(Loc, PD.getDiagID()) << PD;
11976+
const PartialDiagnostic &PD,
11977+
FunctionDecl *FD = nullptr) {
11978+
return targetDiag(Loc, PD.getDiagID(), FD) << PD;
1197711979
}
1197811980

1197911981
/// Check if the expression is allowed to be used in expressions for the
1198011982
/// offloading devices.
11981-
void checkDeviceDecl(const ValueDecl *D, SourceLocation Loc);
11983+
void checkDeviceDecl(ValueDecl *D, SourceLocation Loc);
1198211984

1198311985
enum CUDAFunctionTarget {
1198411986
CFT_Device,

‎clang/lib/Sema/Sema.cpp

Lines changed: 22 additions & 14 deletions
Original file line numberDiff line numberDiff line change
@@ -14,6 +14,7 @@
1414
#include "UsedDeclVisitor.h"
1515
#include "clang/AST/ASTContext.h"
1616
#include "clang/AST/ASTDiagnostic.h"
17+
#include "clang/AST/Decl.h"
1718
#include "clang/AST/DeclCXX.h"
1819
#include "clang/AST/DeclFriend.h"
1920
#include "clang/AST/DeclObjC.h"
@@ -1740,11 +1741,12 @@ Sema::SemaDiagnosticBuilder::~SemaDiagnosticBuilder() {
17401741
}
17411742
}
17421743

1743-
Sema::SemaDiagnosticBuilder Sema::targetDiag(SourceLocation Loc,
1744-
unsigned DiagID) {
1744+
Sema::SemaDiagnosticBuilder
1745+
Sema::targetDiag(SourceLocation Loc, unsigned DiagID, FunctionDecl *FD) {
1746+
FD = FD ? FD : getCurFunctionDecl();
17451747
if (LangOpts.OpenMP)
1746-
return LangOpts.OpenMPIsDevice ? diagIfOpenMPDeviceCode(Loc, DiagID)
1747-
: diagIfOpenMPHostCode(Loc, DiagID);
1748+
return LangOpts.OpenMPIsDevice ? diagIfOpenMPDeviceCode(Loc, DiagID, FD)
1749+
: diagIfOpenMPHostCode(Loc, DiagID, FD);
17481750
if (getLangOpts().CUDA)
17491751
return getLangOpts().CUDAIsDevice ? CUDADiagIfDeviceCode(Loc, DiagID)
17501752
: CUDADiagIfHostCode(Loc, DiagID);
@@ -1753,7 +1755,7 @@ Sema::SemaDiagnosticBuilder Sema::targetDiag(SourceLocation Loc,
17531755
return SYCLDiagIfDeviceCode(Loc, DiagID);
17541756

17551757
return SemaDiagnosticBuilder(SemaDiagnosticBuilder::K_Immediate, Loc, DiagID,
1756-
getCurFunctionDecl(), *this);
1758+
FD, *this);
17571759
}
17581760

17591761
Sema::SemaDiagnosticBuilder Sema::Diag(SourceLocation Loc, unsigned DiagID,
@@ -1772,15 +1774,14 @@ Sema::SemaDiagnosticBuilder Sema::Diag(SourceLocation Loc, unsigned DiagID,
17721774
DiagID, getCurFunctionDecl(), *this);
17731775
}
17741776

1775-
SemaDiagnosticBuilder DB =
1776-
getLangOpts().CUDAIsDevice
1777-
? CUDADiagIfDeviceCode(Loc, DiagID)
1778-
: CUDADiagIfHostCode(Loc, DiagID);
1777+
SemaDiagnosticBuilder DB = getLangOpts().CUDAIsDevice
1778+
? CUDADiagIfDeviceCode(Loc, DiagID)
1779+
: CUDADiagIfHostCode(Loc, DiagID);
17791780
SetIsLastErrorImmediate(DB.isImmediate());
17801781
return DB;
17811782
}
17821783

1783-
void Sema::checkDeviceDecl(const ValueDecl *D, SourceLocation Loc) {
1784+
void Sema::checkDeviceDecl(ValueDecl *D, SourceLocation Loc) {
17841785
if (isUnevaluatedContext())
17851786
return;
17861787

@@ -1798,13 +1799,17 @@ void Sema::checkDeviceDecl(const ValueDecl *D, SourceLocation Loc) {
17981799
return;
17991800
}
18001801

1802+
// Try to associate errors with the lexical context, if that is a function, or
1803+
// the value declaration otherwise.
1804+
FunctionDecl *FD =
1805+
isa<FunctionDecl>(C) ? cast<FunctionDecl>(C) : dyn_cast<FunctionDecl>(D);
18011806
auto CheckType = [&](QualType Ty) {
18021807
if (Ty->isDependentType())
18031808
return;
18041809

18051810
if (Ty->isExtIntType()) {
18061811
if (!Context.getTargetInfo().hasExtIntType()) {
1807-
targetDiag(Loc, diag::err_device_unsupported_type)
1812+
targetDiag(Loc, diag::err_device_unsupported_type, FD)
18081813
<< D << false /*show bit size*/ << 0 /*bitsize*/
18091814
<< Ty << Context.getTargetInfo().getTriple().str();
18101815
}
@@ -1817,11 +1822,12 @@ void Sema::checkDeviceDecl(const ValueDecl *D, SourceLocation Loc) {
18171822
!Context.getTargetInfo().hasFloat128Type()) ||
18181823
(Ty->isIntegerType() && Context.getTypeSize(Ty) == 128 &&
18191824
!Context.getTargetInfo().hasInt128Type())) {
1820-
targetDiag(Loc, diag::err_device_unsupported_type)
1825+
if (targetDiag(Loc, diag::err_device_unsupported_type, FD)
18211826
<< D << true /*show bit size*/
18221827
<< static_cast<unsigned>(Context.getTypeSize(Ty)) << Ty
1823-
<< Context.getTargetInfo().getTriple().str();
1824-
targetDiag(D->getLocation(), diag::note_defined_here) << D;
1828+
<< Context.getTargetInfo().getTriple().str())
1829+
D->setInvalidDecl();
1830+
targetDiag(D->getLocation(), diag::note_defined_here, FD) << D;
18251831
}
18261832
};
18271833

@@ -1833,6 +1839,8 @@ void Sema::checkDeviceDecl(const ValueDecl *D, SourceLocation Loc) {
18331839
CheckType(ParamTy);
18341840
CheckType(FPTy->getReturnType());
18351841
}
1842+
if (const auto *FNPTy = dyn_cast<FunctionNoProtoType>(Ty))
1843+
CheckType(FNPTy->getReturnType());
18361844
}
18371845

18381846
/// Looks through the macro-expansion chain for the given

‎clang/lib/Sema/SemaDecl.cpp

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -9421,6 +9421,9 @@ Sema::ActOnFunctionDeclarator(Scope *S, Declarator &D, DeclContext *DC,
94219421
}
94229422
}
94239423

9424+
if (LangOpts.SYCLIsDevice || (LangOpts.OpenMP && LangOpts.OpenMPIsDevice))
9425+
checkDeviceDecl(NewFD, D.getBeginLoc());
9426+
94249427
if (!getLangOpts().CPlusPlus) {
94259428
// Perform semantic checking on the function declaration.
94269429
if (!NewFD->isInvalidDecl() && NewFD->isMain())

‎clang/lib/Sema/SemaExpr.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -373,7 +373,7 @@ bool Sema::DiagnoseUseOfDecl(NamedDecl *D, ArrayRef<SourceLocation> Locs,
373373
}
374374

375375
if (LangOpts.SYCLIsDevice || (LangOpts.OpenMP && LangOpts.OpenMPIsDevice)) {
376-
if (const auto *VD = dyn_cast<ValueDecl>(D))
376+
if (auto *VD = dyn_cast<ValueDecl>(D))
377377
checkDeviceDecl(VD, Loc);
378378

379379
if (!Context.getTargetInfo().isTLSSupported())

‎clang/lib/Sema/SemaOpenMP.cpp

Lines changed: 7 additions & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -1898,11 +1898,11 @@ enum class FunctionEmissionStatus {
18981898
} // anonymous namespace
18991899

19001900
Sema::SemaDiagnosticBuilder Sema::diagIfOpenMPDeviceCode(SourceLocation Loc,
1901-
unsigned DiagID) {
1901+
unsigned DiagID,
1902+
FunctionDecl *FD) {
19021903
assert(LangOpts.OpenMP && LangOpts.OpenMPIsDevice &&
19031904
"Expected OpenMP device compilation.");
19041905

1905-
FunctionDecl *FD = getCurFunctionDecl();
19061906
SemaDiagnosticBuilder::Kind Kind = SemaDiagnosticBuilder::K_Nop;
19071907
if (FD) {
19081908
FunctionEmissionStatus FES = getEmissionStatus(FD);
@@ -1925,14 +1925,15 @@ Sema::SemaDiagnosticBuilder Sema::diagIfOpenMPDeviceCode(SourceLocation Loc,
19251925
}
19261926
}
19271927

1928-
return SemaDiagnosticBuilder(Kind, Loc, DiagID, getCurFunctionDecl(), *this);
1928+
return SemaDiagnosticBuilder(Kind, Loc, DiagID, FD, *this);
19291929
}
19301930

19311931
Sema::SemaDiagnosticBuilder Sema::diagIfOpenMPHostCode(SourceLocation Loc,
1932-
unsigned DiagID) {
1932+
unsigned DiagID,
1933+
FunctionDecl *FD) {
19331934
assert(LangOpts.OpenMP && !LangOpts.OpenMPIsDevice &&
19341935
"Expected OpenMP host compilation.");
1935-
FunctionEmissionStatus FES = getEmissionStatus(getCurFunctionDecl());
1936+
FunctionEmissionStatus FES = getEmissionStatus(FD);
19361937
SemaDiagnosticBuilder::Kind Kind = SemaDiagnosticBuilder::K_Nop;
19371938
switch (FES) {
19381939
case FunctionEmissionStatus::Emitted:
@@ -1948,7 +1949,7 @@ Sema::SemaDiagnosticBuilder Sema::diagIfOpenMPHostCode(SourceLocation Loc,
19481949
break;
19491950
}
19501951

1951-
return SemaDiagnosticBuilder(Kind, Loc, DiagID, getCurFunctionDecl(), *this);
1952+
return SemaDiagnosticBuilder(Kind, Loc, DiagID, FD, *this);
19521953
}
19531954

19541955
static OpenMPDefaultmapClauseKind

‎clang/test/OpenMP/nvptx_unsupported_type_messages.cpp

Lines changed: 45 additions & 20 deletions
Original file line numberDiff line numberDiff line change
@@ -39,10 +39,12 @@ struct T1 {
3939
};
4040

4141
#ifndef _ARCH_PPC
42-
// expected-note@+1 {{'boo' defined here}}
42+
// expected-error@+2 {{'boo' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
43+
// expected-note@+1 2{{'boo' defined here}}
4344
void boo(__float128 A) { return; }
4445
#else
45-
// expected-note@+1 {{'boo' defined here}}
46+
// expected-error@+2 {{'boo' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
47+
// expected-note@+1 2{{'boo' defined here}}
4648
void boo(long double A) { return; }
4749
#endif
4850
#pragma omp declare target
@@ -51,10 +53,11 @@ T f = a;
5153
void foo(T a = T()) {
5254
a = a + f; // expected-note {{called by 'foo'}}
5355
#ifndef _ARCH_PPC
54-
// expected-error@+4 {{'boo' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
56+
// expected-error@+5 {{'boo' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
5557
#else
56-
// expected-error@+2 {{'boo' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
58+
// expected-error@+3 {{'boo' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
5759
#endif
60+
// expected-note@+1 {{called by 'foo'}}
5861
boo(0);
5962
return;
6063
}
@@ -98,28 +101,49 @@ void dead_template_declare_target() {
98101
a = &b;
99102
}
100103

101-
// TODO: We should diagnose the return type and argument type here.
104+
// expected-note@+2 {{'ld_return1a' defined here}}
105+
// expected-error@+1 {{'ld_return1a' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
102106
long double ld_return1a() { return 0; }
107+
// expected-note@+2 {{'ld_arg1a' defined here}}
108+
// expected-error@+1 {{'ld_arg1a' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
103109
void ld_arg1a(long double ld) {}
104110

105111
// TODO: We should diagnose the return type and argument type here.
106112
typedef long double ld_ty;
113+
// expected-note@+2 {{'ld_return1b' defined here}}
114+
// expected-error@+1 {{'ld_return1b' requires 128 bit size 'ld_ty' (aka 'long double') type support, but device 'nvptx64-unknown-unknown' does not support it}}
107115
ld_ty ld_return1b() { return 0; }
116+
// expected-note@+2 {{'ld_arg1b' defined here}}
117+
// expected-error@+1 {{'ld_arg1b' requires 128 bit size 'ld_ty' (aka 'long double') type support, but device 'nvptx64-unknown-unknown' does not support it}}
108118
void ld_arg1b(ld_ty ld) {}
109119

120+
// TODO: These errors should not be emitted.
121+
// expected-note@+2 {{'ld_return1c' defined here}}
122+
// expected-error@+1 {{'ld_return1c' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
110123
static long double ld_return1c() { return 0; }
124+
// expected-note@+2 {{'ld_arg1c' defined here}}
125+
// expected-error@+1 {{'ld_arg1c' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
111126
static void ld_arg1c(long double ld) {}
112127

128+
// TODO: These errors should not be emitted.
129+
// expected-note@+2 {{'ld_return1d' defined here}}
130+
// expected-error@+1 {{'ld_return1d' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
113131
inline long double ld_return1d() { return 0; }
132+
// expected-note@+2 {{'ld_arg1d' defined here}}
133+
// expected-error@+1 {{'ld_arg1d' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
114134
inline void ld_arg1d(long double ld) {}
115135

136+
// expected-error@+2 {{'ld_return1e' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
116137
// expected-note@+1 {{'ld_return1e' defined here}}
117138
static long double ld_return1e() { return 0; }
139+
// expected-error@+2 {{'ld_arg1e' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
118140
// expected-note@+1 {{'ld_arg1e' defined here}}
119141
static void ld_arg1e(long double ld) {}
120142

143+
// expected-error@+2 {{'ld_return1f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
121144
// expected-note@+1 {{'ld_return1f' defined here}}
122145
inline long double ld_return1f() { return 0; }
146+
// expected-error@+2 {{'ld_arg1f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
123147
// expected-note@+1 {{'ld_arg1f' defined here}}
124148
inline void ld_arg1f(long double ld) {}
125149

@@ -152,46 +176,47 @@ static void ld_use4() {
152176
}
153177

154178
void external() {
155-
// expected-error@+1 {{'ld_return1e' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
156179
void *p1 = reinterpret_cast<void*>(&ld_return1e);
157-
// expected-error@+1 {{'ld_arg1e' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
158180
void *p2 = reinterpret_cast<void*>(&ld_arg1e);
159-
// expected-error@+1 {{'ld_return1f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
160181
void *p3 = reinterpret_cast<void*>(&ld_return1f);
161-
// expected-error@+1 {{'ld_arg1f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
162182
void *p4 = reinterpret_cast<void*>(&ld_arg1f);
163183
void *p5 = reinterpret_cast<void*>(&ld_use3);
164184
void *p6 = reinterpret_cast<void*>(&ld_use4);
165185
}
166186

167187
#ifndef _ARCH_PPC
168-
// TODO: We should diagnose the return type and argument type here.
188+
// expected-note@+2 {{'ld_return2a' defined here}}
189+
// expected-error@+1 {{'ld_return2a' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
169190
__float128 ld_return2a() { return 0; }
191+
// expected-note@+2 {{'ld_arg2a' defined here}}
192+
// expected-error@+1 {{'ld_arg2a' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
170193
void ld_arg2a(__float128 ld) {}
171194

172-
// TODO: We should diagnose the return type and argument type here.
173195
typedef __float128 fp128_ty;
196+
// expected-note@+2 {{'ld_return2b' defined here}}
197+
// expected-error@+1 {{'ld_return2b' requires 128 bit size 'fp128_ty' (aka '__float128') type support, but device 'nvptx64-unknown-unknown' does not support it}}
174198
fp128_ty ld_return2b() { return 0; }
199+
// expected-note@+2 {{'ld_arg2b' defined here}}
200+
// expected-error@+1 {{'ld_arg2b' requires 128 bit size 'fp128_ty' (aka '__float128') type support, but device 'nvptx64-unknown-unknown' does not support it}}
175201
void ld_arg2b(fp128_ty ld) {}
176202
#endif
177203

178204
#pragma omp end declare target
179205

180206
// TODO: There should not be an error here, dead_inline is never emitted.
181-
// expected-note@+1 3{{'f' defined here}}
207+
// expected-note@+1 {{'f' defined here}}
182208
inline long double dead_inline(long double f) {
183209
#pragma omp target map(f)
184-
// TODO: We should not emit the same error message 3 times, here and elsewhere in this file.
185-
// expected-error@+1 3{{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
210+
// expected-error@+1 {{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
186211
f = 1;
187212
return f;
188213
}
189214

190215
// TODO: There should not be an error here, dead_static is never emitted.
191-
// expected-note@+1 3{{'f' defined here}}
216+
// expected-note@+1 {{'f' defined here}}
192217
static long double dead_static(long double f) {
193218
#pragma omp target map(f)
194-
// expected-error@+1 3{{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
219+
// expected-error@+1 {{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
195220
f = 1;
196221
return f;
197222
}
@@ -204,18 +229,18 @@ long double dead_template(long double f) {
204229
}
205230

206231
#ifndef _ARCH_PPC
207-
// expected-note@+1 3{{'f' defined here}}
232+
// expected-note@+1 {{'f' defined here}}
208233
__float128 foo2(__float128 f) {
209234
#pragma omp target map(f)
210-
// expected-error@+1 3{{'f' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
235+
// expected-error@+1 {{'f' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
211236
f = 1;
212237
return f;
213238
}
214239
#else
215-
// expected-note@+1 3{{'f' defined here}}
240+
// expected-note@+1 {{'f' defined here}}
216241
long double foo3(long double f) {
217242
#pragma omp target map(f)
218-
// expected-error@+1 3{{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
243+
// expected-error@+1 {{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
219244
f = 1;
220245
return f;
221246
}

0 commit comments

Comments
 (0)