[clang] 3dbcea8 - Reland [clang] Check unsupported types in expressions
Andrew Savonichev via cfe-commits
cfe-commits at lists.llvm.org
Fri Oct 15 03:55:59 PDT 2021
Author: Andrew Savonichev
Date: 2021-10-15T13:55:36+03:00
New Revision: 3dbcea8b957a5f0094f1d5a73b7a8ebebeedebb6
URL: https://github.com/llvm/llvm-project/commit/3dbcea8b957a5f0094f1d5a73b7a8ebebeedebb6
DIFF: https://github.com/llvm/llvm-project/commit/3dbcea8b957a5f0094f1d5a73b7a8ebebeedebb6.diff
LOG: Reland [clang] Check unsupported types in expressions
This was committed as ec6c847179fd, but then reverted after a failure
in: https://lab.llvm.org/buildbot/#/builders/84/builds/13983
I was not able to reproduce the problem, but I added an extra check
for a NULL QualType just in case.
Original comit message:
The patch adds missing diagnostics for cases like:
float F3 = ((__float128)F1 * (__float128)F2) / 2.0f;
Sema::checkDeviceDecl (renamed to checkTypeSupport) is changed to work
with a type without the corresponding ValueDecl. It is also refactored
so that host diagnostics for unsupported types can be added here as
well.
Differential Revision: https://reviews.llvm.org/D109315
Added:
Modified:
clang/include/clang/Basic/DiagnosticSemaKinds.td
clang/include/clang/Sema/Sema.h
clang/lib/Sema/Sema.cpp
clang/lib/Sema/SemaDecl.cpp
clang/lib/Sema/SemaExpr.cpp
clang/test/CodeGen/ibm128-unsupported.c
clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
clang/test/SemaSYCL/float128.cpp
clang/test/SemaSYCL/int128.cpp
Removed:
################################################################################
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index 1634c23db3d26..e0c36961a9124 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -10713,8 +10713,8 @@ def err_omp_invariant_or_linear_dependency : Error<
"expected loop invariant expression or '<invariant1> * %0 + <invariant2>' kind of expression">;
def err_omp_wrong_dependency_iterator_type : Error<
"expected an integer or a pointer type of the outer loop counter '%0' for non-rectangular nests">;
-def err_device_unsupported_type
- : Error<"%0 requires %select{|%2 bit size}1 %3 type support, but device "
+def err_target_unsupported_type
+ : Error<"%0 requires %select{|%2 bit size}1 %3 type support, but target "
"'%4' does not support it">;
def err_omp_lambda_capture_in_declare_target_not_to : Error<
"variable captured in declare target region must appear in a to clause">;
diff --git a/clang/include/clang/Sema/Sema.h b/clang/include/clang/Sema/Sema.h
index ff759cc77ea53..4d5f9919ce9e6 100644
--- a/clang/include/clang/Sema/Sema.h
+++ b/clang/include/clang/Sema/Sema.h
@@ -12219,9 +12219,9 @@ class Sema final {
return targetDiag(Loc, PD.getDiagID(), FD) << PD;
}
- /// Check if the expression is allowed to be used in expressions for the
- /// offloading devices.
- void checkDeviceDecl(ValueDecl *D, SourceLocation Loc);
+ /// Check if the type is allowed to be used for the current target.
+ void checkTypeSupport(QualType Ty, SourceLocation Loc,
+ ValueDecl *D = nullptr);
enum CUDAFunctionTarget {
CFT_Device,
diff --git a/clang/lib/Sema/Sema.cpp b/clang/lib/Sema/Sema.cpp
index 88833f233adc1..9fae2243cbbf6 100644
--- a/clang/lib/Sema/Sema.cpp
+++ b/clang/lib/Sema/Sema.cpp
@@ -1854,8 +1854,11 @@ Sema::SemaDiagnosticBuilder Sema::Diag(SourceLocation Loc, unsigned DiagID,
return DB;
}
-void Sema::checkDeviceDecl(ValueDecl *D, SourceLocation Loc) {
- if (isUnevaluatedContext())
+void Sema::checkTypeSupport(QualType Ty, SourceLocation Loc, ValueDecl *D) {
+ if (!LangOpts.SYCLIsDevice && !(LangOpts.OpenMP && LangOpts.OpenMPIsDevice))
+ return;
+
+ if (isUnevaluatedContext() || Ty.isNull())
return;
Decl *C = cast<Decl>(getCurLexicalContext());
@@ -1874,16 +1877,22 @@ void Sema::checkDeviceDecl(ValueDecl *D, SourceLocation Loc) {
// Try to associate errors with the lexical context, if that is a function, or
// the value declaration otherwise.
- FunctionDecl *FD =
- isa<FunctionDecl>(C) ? cast<FunctionDecl>(C) : dyn_cast<FunctionDecl>(D);
+ FunctionDecl *FD = isa<FunctionDecl>(C) ? cast<FunctionDecl>(C)
+ : dyn_cast_or_null<FunctionDecl>(D);
+
auto CheckType = [&](QualType Ty) {
if (Ty->isDependentType())
return;
if (Ty->isExtIntType()) {
if (!Context.getTargetInfo().hasExtIntType()) {
- targetDiag(Loc, diag::err_device_unsupported_type, FD)
- << D << false /*show bit size*/ << 0 /*bitsize*/
+ PartialDiagnostic PD = PDiag(diag::err_target_unsupported_type);
+ if (D)
+ PD << D;
+ else
+ PD << "expression";
+ targetDiag(Loc, PD, FD)
+ << false /*show bit size*/ << 0 /*bitsize*/
<< Ty << Context.getTargetInfo().getTriple().str();
}
return;
@@ -1907,16 +1916,24 @@ void Sema::checkDeviceDecl(ValueDecl *D, SourceLocation Loc) {
(Ty->isIntegerType() && Context.getTypeSize(Ty) == 128 &&
!Context.getTargetInfo().hasInt128Type()) ||
LongDoubleMismatched) {
- if (targetDiag(Loc, diag::err_device_unsupported_type, FD)
- << D << true /*show bit size*/
+ PartialDiagnostic PD = PDiag(diag::err_target_unsupported_type);
+ if (D)
+ PD << D;
+ else
+ PD << "expression";
+
+ if (targetDiag(Loc, PD, FD)
+ << true /*show bit size*/
<< static_cast<unsigned>(Context.getTypeSize(Ty)) << Ty
- << Context.getTargetInfo().getTriple().str())
- D->setInvalidDecl();
- targetDiag(D->getLocation(), diag::note_defined_here, FD) << D;
+ << Context.getTargetInfo().getTriple().str()) {
+ if (D)
+ D->setInvalidDecl();
+ }
+ if (D)
+ targetDiag(D->getLocation(), diag::note_defined_here, FD) << D;
}
};
- QualType Ty = D->getType();
CheckType(Ty);
if (const auto *FPTy = dyn_cast<FunctionProtoType>(Ty)) {
diff --git a/clang/lib/Sema/SemaDecl.cpp b/clang/lib/Sema/SemaDecl.cpp
index 7fff85a360d67..1491407b0533c 100644
--- a/clang/lib/Sema/SemaDecl.cpp
+++ b/clang/lib/Sema/SemaDecl.cpp
@@ -9570,8 +9570,7 @@ Sema::ActOnFunctionDeclarator(Scope *S, Declarator &D, DeclContext *DC,
}
}
- if (LangOpts.SYCLIsDevice || (LangOpts.OpenMP && LangOpts.OpenMPIsDevice))
- checkDeviceDecl(NewFD, D.getBeginLoc());
+ checkTypeSupport(NewFD->getType(), D.getBeginLoc(), NewFD);
if (!getLangOpts().CPlusPlus) {
// Perform semantic checking on the function declaration.
diff --git a/clang/lib/Sema/SemaExpr.cpp b/clang/lib/Sema/SemaExpr.cpp
index dc2455e0afbcd..94b44714b530d 100644
--- a/clang/lib/Sema/SemaExpr.cpp
+++ b/clang/lib/Sema/SemaExpr.cpp
@@ -368,10 +368,10 @@ bool Sema::DiagnoseUseOfDecl(NamedDecl *D, ArrayRef<SourceLocation> Locs,
diagnoseUseOfInternalDeclInInlineFunction(*this, D, Loc);
- if (LangOpts.SYCLIsDevice || (LangOpts.OpenMP && LangOpts.OpenMPIsDevice)) {
- if (auto *VD = dyn_cast<ValueDecl>(D))
- checkDeviceDecl(VD, Loc);
+ if (auto *VD = dyn_cast<ValueDecl>(D))
+ checkTypeSupport(VD->getType(), Loc, VD);
+ if (LangOpts.SYCLIsDevice || (LangOpts.OpenMP && LangOpts.OpenMPIsDevice)) {
if (!Context.getTargetInfo().isTLSSupported())
if (const auto *VD = dyn_cast<VarDecl>(D))
if (VD->getTLSKind() != VarDecl::TLS_None)
@@ -14165,6 +14165,9 @@ ExprResult Sema::CreateBuiltinBinOp(SourceLocation OpLoc,
}
}
+ checkTypeSupport(LHSExpr->getType(), OpLoc, /*ValueDecl*/ nullptr);
+ checkTypeSupport(RHSExpr->getType(), OpLoc, /*ValueDecl*/ nullptr);
+
switch (Opc) {
case BO_Assign:
ResultTy = CheckAssignmentOperands(LHS.get(), RHS, OpLoc, QualType());
diff --git a/clang/test/CodeGen/ibm128-unsupported.c b/clang/test/CodeGen/ibm128-unsupported.c
index 4ade7ddaa1e39..4c4930f3a7ce4 100644
--- a/clang/test/CodeGen/ibm128-unsupported.c
+++ b/clang/test/CodeGen/ibm128-unsupported.c
@@ -9,7 +9,7 @@ void foo(__ibm128 x); // expected-note {{'foo' defined here}}
void loop(int n, __ibm128 *arr) {
#pragma omp target parallel
for (int i = 0; i < n; ++i) {
- // expected-error at +1 {{'foo' requires 128 bit size '__ibm128' type support, but device 'x86_64' does not support it}}
+ // expected-error at +1 {{'foo' requires 128 bit size '__ibm128' type support, but target 'x86_64' does not support it}}
foo(arr[i]);
}
}
diff --git a/clang/test/OpenMP/nvptx_unsupported_type_messages.cpp b/clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
index a319c78f73c56..e57ef1d288a32 100644
--- a/clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
+++ b/clang/test/OpenMP/nvptx_unsupported_type_messages.cpp
@@ -16,9 +16,11 @@ struct T {
char c;
T() : a(12), f(15) {}
#ifndef _ARCH_PPC
-// expected-error at +5 {{'f' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +7 {{'f' requires 128 bit size '__float128' type support, but target 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +6 {{expression requires 128 bit size '__float128' type support, but target 'nvptx64-unknown-unknown' does not support it}}
#else
-// expected-error at +3 {{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +4 {{'f' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +3 {{expression requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
#endif
T &operator+(T &b) {
f += b.a;
@@ -39,11 +41,11 @@ struct T1 {
};
#ifndef _ARCH_PPC
-// expected-error at +2 {{'boo' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +2 {{'boo' requires 128 bit size '__float128' type support, but target 'nvptx64-unknown-unknown' does not support it}}
// expected-note at +1 2{{'boo' defined here}}
void boo(__float128 A) { return; }
#else
-// expected-error at +2 {{'boo' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +2 {{'boo' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
// expected-note at +1 2{{'boo' defined here}}
void boo(long double A) { return; }
#endif
@@ -53,9 +55,9 @@ T f = a;
void foo(T a = T()) {
a = a + f; // expected-note {{called by 'foo'}}
#ifndef _ARCH_PPC
-// expected-error at +5 {{'boo' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +5 {{'boo' requires 128 bit size '__float128' type support, but target 'nvptx64-unknown-unknown' does not support it}}
#else
-// expected-error at +3 {{'boo' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +3 {{'boo' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
#endif
// expected-note at +1 {{called by 'foo'}}
boo(0);
@@ -96,18 +98,18 @@ void dead_template_declare_target() {
}
// expected-note at +2 {{'ld_return1a' defined here}}
-// expected-error at +1 {{'ld_return1a' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_return1a' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
long double ld_return1a() { return 0; }
// expected-note at +2 {{'ld_arg1a' defined here}}
-// expected-error at +1 {{'ld_arg1a' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_arg1a' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
void ld_arg1a(long double ld) {}
typedef long double ld_ty;
// expected-note at +2 {{'ld_return1b' defined here}}
-// expected-error at +1 {{'ld_return1b' requires 128 bit size 'ld_ty' (aka 'long double') type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_return1b' requires 128 bit size 'ld_ty' (aka 'long double') type support, but target 'nvptx64-unknown-unknown' does not support it}}
ld_ty ld_return1b() { return 0; }
// expected-note at +2 {{'ld_arg1b' defined here}}
-// expected-error at +1 {{'ld_arg1b' requires 128 bit size 'ld_ty' (aka 'long double') type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_arg1b' requires 128 bit size 'ld_ty' (aka 'long double') type support, but target 'nvptx64-unknown-unknown' does not support it}}
void ld_arg1b(ld_ty ld) {}
static long double ld_return1c() { return 0; }
@@ -138,24 +140,26 @@ static void ld_use2() {
inline void ld_use3() {
// expected-note at +1 {{'ld' defined here}}
long double ld = 0;
-// expected-error at +1 {{'ld' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +2 {{expression requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
ld += 1;
}
static void ld_use4() {
// expected-note at +1 {{'ld' defined here}}
long double ld = 0;
-// expected-error at +1 {{'ld' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +2 {{expression requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
ld += 1;
}
void external() {
-// expected-error at +1 {{'ld_return1e' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_return1e' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
void *p1 = reinterpret_cast<void*>(&ld_return1e);
-// expected-error at +1 {{'ld_arg1e' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_arg1e' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
void *p2 = reinterpret_cast<void*>(&ld_arg1e);
-// expected-error at +1 {{'ld_return1f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_return1f' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
void *p3 = reinterpret_cast<void*>(&ld_return1f);
-// expected-error at +1 {{'ld_arg1f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_arg1f' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
void *p4 = reinterpret_cast<void*>(&ld_arg1f);
// TODO: The error message "called by" is not great.
// expected-note at +1 {{called by 'external'}}
@@ -166,18 +170,18 @@ void external() {
#ifndef _ARCH_PPC
// expected-note at +2 {{'ld_return2a' defined here}}
-// expected-error at +1 {{'ld_return2a' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_return2a' requires 128 bit size '__float128' type support, but target 'nvptx64-unknown-unknown' does not support it}}
__float128 ld_return2a() { return 0; }
// expected-note at +2 {{'ld_arg2a' defined here}}
-// expected-error at +1 {{'ld_arg2a' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_arg2a' requires 128 bit size '__float128' type support, but target 'nvptx64-unknown-unknown' does not support it}}
void ld_arg2a(__float128 ld) {}
typedef __float128 fp128_ty;
// expected-note at +2 {{'ld_return2b' defined here}}
-// expected-error at +1 {{'ld_return2b' requires 128 bit size 'fp128_ty' (aka '__float128') type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_return2b' requires 128 bit size 'fp128_ty' (aka '__float128') type support, but target 'nvptx64-unknown-unknown' does not support it}}
fp128_ty ld_return2b() { return 0; }
// expected-note at +2 {{'ld_arg2b' defined here}}
-// expected-error at +1 {{'ld_arg2b' requires 128 bit size 'fp128_ty' (aka '__float128') type support, but device 'nvptx64-unknown-unknown' does not support it}}
+// expected-error at +1 {{'ld_arg2b' requires 128 bit size 'fp128_ty' (aka '__float128') type support, but target 'nvptx64-unknown-unknown' does not support it}}
void ld_arg2b(fp128_ty ld) {}
#endif
@@ -187,7 +191,7 @@ void ld_arg2b(fp128_ty ld) {}
// expected-note at +1 {{'f' defined here}}
inline long double dead_inline(long double f) {
#pragma omp target map(f)
- // expected-error at +1 {{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+ // expected-error at +1 {{'f' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
f = 1;
return f;
}
@@ -196,7 +200,7 @@ inline long double dead_inline(long double f) {
// expected-note at +1 {{'f' defined here}}
static long double dead_static(long double f) {
#pragma omp target map(f)
- // expected-error at +1 {{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+ // expected-error at +1 {{'f' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
f = 1;
return f;
}
@@ -212,7 +216,7 @@ long double dead_template(long double f) {
// expected-note at +1 {{'f' defined here}}
__float128 foo2(__float128 f) {
#pragma omp target map(f)
- // expected-error at +1 {{'f' requires 128 bit size '__float128' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+ // expected-error at +1 {{'f' requires 128 bit size '__float128' type support, but target 'nvptx64-unknown-unknown' does not support it}}
f = 1;
return f;
}
@@ -220,7 +224,7 @@ __float128 foo2(__float128 f) {
// expected-note at +1 {{'f' defined here}}
long double foo3(long double f) {
#pragma omp target map(f)
- // expected-error at +1 {{'f' requires 128 bit size 'long double' type support, but device 'nvptx64-unknown-unknown' does not support it}}
+ // expected-error at +1 {{'f' requires 128 bit size 'long double' type support, but target 'nvptx64-unknown-unknown' does not support it}}
f = 1;
return f;
}
diff --git a/clang/test/SemaSYCL/float128.cpp b/clang/test/SemaSYCL/float128.cpp
index b91535eda4897..4629437afa4e2 100644
--- a/clang/test/SemaSYCL/float128.cpp
+++ b/clang/test/SemaSYCL/float128.cpp
@@ -26,20 +26,28 @@ void usage() {
// expected-note at +1 3{{'A' defined here}}
__float128 A;
Z<__float128> C;
- // expected-error at +2 {{'A' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
- // expected-error at +1 {{'field1' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
+ // expected-error at +3 2{{expression requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
+ // expected-error at +2 {{'A' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'field1' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
C.field1 = A;
- // expected-error at +1 {{'bigfield' requires 128 bit size 'Z::BIGTYPE' (aka '__float128') type support, but device 'spir64' does not support it}}
+ // expected-error at +2 {{expression requires 128 bit size 'Z::BIGTYPE' (aka '__float128') type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'bigfield' requires 128 bit size 'Z::BIGTYPE' (aka '__float128') type support, but target 'spir64' does not support it}}
C.bigfield += 1.0;
- // expected-error at +1 {{'A' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
+ // expected-error at +1 {{'A' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
auto foo1 = [=]() {
__float128 AA;
// expected-note at +2 {{'BB' defined here}}
- // expected-error at +1 {{'A' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
+ // expected-error at +1 {{'A' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
auto BB = A;
- // expected-error at +1 {{'BB' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
+ // expected-error at +2 {{expression requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'BB' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
BB += 1;
+
+ float F1 = 0.1f;
+ float F2 = 0.1f;
+ // expected-error at +1 3{{expression requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
+ float F3 = ((__float128)F1 * (__float128)F2) / 2.0f;
};
// expected-note at +1 {{called by 'usage'}}
@@ -50,7 +58,7 @@ template <typename t>
void foo2(){};
// expected-note at +3 {{'P' defined here}}
-// expected-error at +2 {{'P' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
+// expected-error at +2 {{'P' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
// expected-note at +1 2{{'foo' defined here}}
__float128 foo(__float128 P) { return P; }
@@ -66,12 +74,14 @@ int main() {
host_ok();
kernel<class variables>([=]() {
decltype(CapturedToDevice) D;
- // expected-error at +1 {{'CapturedToDevice' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
+ // expected-error at +1 {{'CapturedToDevice' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
auto C = CapturedToDevice;
Z<__float128> S;
- // expected-error at +1 {{'field1' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
+ // expected-error at +2 {{expression requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'field1' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
S.field1 += 1;
- // expected-error at +1 {{'field' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
+ // expected-error at +2 {{expression requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'field' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
S.field = 1;
});
@@ -81,8 +91,8 @@ int main() {
// expected-note at +1 {{'BBBB' defined here}}
BIGTY BBBB;
// expected-note at +3 {{called by 'operator()'}}
- // expected-error at +2 2{{'foo' requires 128 bit size '__float128' type support, but device 'spir64' does not support it}}
- // expected-error at +1 {{'BBBB' requires 128 bit size 'BIGTY' (aka '__float128') type support, but device 'spir64' does not support it}}
+ // expected-error at +2 {{'BBBB' requires 128 bit size 'BIGTY' (aka '__float128') type support, but target 'spir64' does not support it}}
+ // expected-error at +1 2{{'foo' requires 128 bit size '__float128' type support, but target 'spir64' does not support it}}
auto A = foo(BBBB);
});
diff --git a/clang/test/SemaSYCL/int128.cpp b/clang/test/SemaSYCL/int128.cpp
index f6f92c237a9c5..bfa9a04188027 100644
--- a/clang/test/SemaSYCL/int128.cpp
+++ b/clang/test/SemaSYCL/int128.cpp
@@ -26,19 +26,22 @@ void usage() {
// expected-note at +1 3{{'A' defined here}}
__int128 A;
Z<__int128> C;
- // expected-error at +2 {{'A' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
- // expected-error at +1 {{'field1' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+ // expected-error at +3 2{{expression requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
+ // expected-error at +2 {{'A' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'field1' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
C.field1 = A;
- // expected-error at +1 {{'bigfield' requires 128 bit size 'Z::BIGTYPE' (aka '__int128') type support, but device 'spir64' does not support it}}
+ // expected-error at +2 {{expression requires 128 bit size 'Z::BIGTYPE' (aka '__int128') type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'bigfield' requires 128 bit size 'Z::BIGTYPE' (aka '__int128') type support, but target 'spir64' does not support it}}
C.bigfield += 1.0;
- // expected-error at +1 {{'A' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+ // expected-error at +1 {{'A' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
auto foo1 = [=]() {
__int128 AA;
// expected-note at +2 {{'BB' defined here}}
- // expected-error at +1 {{'A' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+ // expected-error at +1 {{'A' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
auto BB = A;
- // expected-error at +1 {{'BB' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+ // expected-error at +2 {{expression requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'BB' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
BB += 1;
};
@@ -50,7 +53,7 @@ template <typename t>
void foo2(){};
// expected-note at +3 {{'P' defined here}}
-// expected-error at +2 {{'P' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+// expected-error at +2 {{'P' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
// expected-note at +1 2{{'foo' defined here}}
__int128 foo(__int128 P) { return P; }
@@ -58,7 +61,8 @@ void foobar() {
// expected-note at +1 {{'operator __int128' defined here}}
struct X { operator __int128() const; } x;
bool a = false;
- // expected-error at +1 {{'operator __int128' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+ // expected-error at +2 2{{expression requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'operator __int128' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
a = x == __int128(0);
}
@@ -74,12 +78,14 @@ int main() {
host_ok();
kernel<class variables>([=]() {
decltype(CapturedToDevice) D;
- // expected-error at +1 {{'CapturedToDevice' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+ // expected-error at +1 {{'CapturedToDevice' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
auto C = CapturedToDevice;
Z<__int128> S;
- // expected-error at +1 {{'field1' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+ // expected-error at +2 {{expression requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'field1' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
S.field1 += 1;
- // expected-error at +1 {{'field' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+ // expected-error at +2 {{expression requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
+ // expected-error at +1 {{'field' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
S.field = 1;
});
@@ -88,8 +94,8 @@ int main() {
usage();
// expected-note at +1 {{'BBBB' defined here}}
BIGTY BBBB;
- // expected-error at +3 {{'BBBB' requires 128 bit size 'BIGTY' (aka 'unsigned __int128') type support, but device 'spir64' does not support it}}
- // expected-error at +2 2{{'foo' requires 128 bit size '__int128' type support, but device 'spir64' does not support it}}
+ // expected-error at +3 {{'BBBB' requires 128 bit size 'BIGTY' (aka 'unsigned __int128') type support, but target 'spir64' does not support it}}
+ // expected-error at +2 2{{'foo' requires 128 bit size '__int128' type support, but target 'spir64' does not support it}}
// expected-note at +1 1{{called by 'operator()'}}
auto A = foo(BBBB);
// expected-note at +1 {{called by 'operator()'}}
More information about the cfe-commits
mailing list