yaxunl updated this revision to Diff 289814.
yaxunl added a comment.
Defer overload resolution diags only if there are wrong-sided candidates.
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D84364/new/
https://reviews.llvm.org/D84364
Files:
clang/include/clang/Basic/Diagnostic.td
clang/include/clang/Basic/DiagnosticAST.h
clang/include/clang/Basic/DiagnosticAnalysis.h
clang/include/clang/Basic/DiagnosticComment.h
clang/include/clang/Basic/DiagnosticCrossTU.h
clang/include/clang/Basic/DiagnosticDriver.h
clang/include/clang/Basic/DiagnosticFrontend.h
clang/include/clang/Basic/DiagnosticIDs.h
clang/include/clang/Basic/DiagnosticLex.h
clang/include/clang/Basic/DiagnosticParse.h
clang/include/clang/Basic/DiagnosticRefactoring.h
clang/include/clang/Basic/DiagnosticSema.h
clang/include/clang/Basic/DiagnosticSemaKinds.td
clang/include/clang/Basic/DiagnosticSerialization.h
clang/include/clang/Basic/LangOptions.def
clang/include/clang/Driver/Options.td
clang/include/clang/Sema/Sema.h
clang/lib/Basic/DiagnosticIDs.cpp
clang/lib/Driver/ToolChains/Cuda.cpp
clang/lib/Driver/ToolChains/HIP.cpp
clang/lib/Frontend/CompilerInvocation.cpp
clang/lib/Sema/AnalysisBasedWarnings.cpp
clang/lib/Sema/Sema.cpp
clang/lib/Sema/SemaAttr.cpp
clang/lib/Sema/SemaCUDA.cpp
clang/lib/Sema/SemaDecl.cpp
clang/lib/Sema/SemaExprObjC.cpp
clang/lib/Sema/SemaOpenMP.cpp
clang/lib/Sema/SemaOverload.cpp
clang/lib/Sema/SemaSYCL.cpp
clang/lib/Sema/SemaStmt.cpp
clang/lib/Sema/SemaStmtAsm.cpp
clang/lib/Sema/SemaTemplateInstantiate.cpp
clang/lib/Sema/SemaTemplateInstantiateDecl.cpp
clang/lib/Sema/SemaTemplateVariadic.cpp
clang/lib/Sema/SemaType.cpp
clang/test/SemaCUDA/deferred-oeverload.cu
clang/test/TableGen/DiagnosticBase.inc
clang/test/TableGen/deferred-diag.td
clang/tools/diagtool/DiagnosticNames.cpp
clang/utils/TableGen/ClangDiagnosticsEmitter.cpp
Index: clang/utils/TableGen/ClangDiagnosticsEmitter.cpp
===================================================================
--- clang/utils/TableGen/ClangDiagnosticsEmitter.cpp
+++ clang/utils/TableGen/ClangDiagnosticsEmitter.cpp
@@ -1294,6 +1294,11 @@
else
OS << ", false";
+ if (R.getValueAsBit("Deferrable"))
+ OS << ", true";
+ else
+ OS << ", false";
+
// Category number.
OS << ", " << CategoryIDs.getID(getDiagnosticCategory(&R, DGParentMap));
OS << ")\n";
Index: clang/tools/diagtool/DiagnosticNames.cpp
===================================================================
--- clang/tools/diagtool/DiagnosticNames.cpp
+++ clang/tools/diagtool/DiagnosticNames.cpp
@@ -28,7 +28,7 @@
// out of sync easily?
static const DiagnosticRecord BuiltinDiagnosticsByID[] = {
#define DIAG(ENUM,CLASS,DEFAULT_MAPPING,DESC,GROUP, \
- SFINAE,NOWERROR,SHOWINSYSHEADER,CATEGORY) \
+ SFINAE,NOWERROR,SHOWINSYSHEADER,DEFER,CATEGORY) \
{ #ENUM, diag::ENUM, STR_SIZE(#ENUM, uint8_t) },
#include "clang/Basic/DiagnosticCommonKinds.inc"
#include "clang/Basic/DiagnosticCrossTUKinds.inc"
Index: clang/test/TableGen/deferred-diag.td
===================================================================
--- /dev/null
+++ clang/test/TableGen/deferred-diag.td
@@ -0,0 +1,27 @@
+// RUN: clang-tblgen -gen-clang-diags-defs -I%S %s -o - 2>&1 | \
+// RUN: FileCheck --strict-whitespace %s
+include "DiagnosticBase.inc"
+
+// Test usage of Deferrable and NonDeferrable in diagnostics.
+
+def test_default : Error<"This error is non-deferrable by default">;
+// CHECK-DAG: DIAG(test_default, {{.*}}SFINAE_SubstitutionFailure, false, true, false, 0)
+
+def test_deferrable : Error<"This error is deferrable">, Deferrable;
+// CHECK-DAG: DIAG(test_deferrable, {{.*}} SFINAE_SubstitutionFailure, false, true, true, 0)
+
+def test_non_deferrable : Error<"This error is non-deferrable">, NonDeferrable;
+// CHECK-DAG: DIAG(test_non_deferrable, {{.*}} SFINAE_SubstitutionFailure, false, true, false, 0)
+
+let Deferrable = 1 in {
+
+def test_let : Error<"This error is deferrable by let">;
+// CHECK-DAG: DIAG(test_let, {{.*}} SFINAE_SubstitutionFailure, false, true, true, 0)
+
+// Make sure TextSubstitution is allowed in the let Deferrable block.
+def textsub : TextSubstitution<"%select{text1|text2}0">;
+
+def test_let2 : Error<"This error is deferrable by let %sub{textsub}0">;
+// CHECK-DAG: DIAG(test_let2, {{.*}} SFINAE_SubstitutionFailure, false, true, true, 0)
+
+}
\ No newline at end of file
Index: clang/test/TableGen/DiagnosticBase.inc
===================================================================
--- clang/test/TableGen/DiagnosticBase.inc
+++ clang/test/TableGen/DiagnosticBase.inc
@@ -45,6 +45,7 @@
// diagnostics
string Component = "";
string CategoryName = "";
+ bit Deferrable = 0;
}
// Diagnostic Categories. These can be applied to groups or individual
@@ -75,6 +76,7 @@
bit AccessControl = 0;
bit WarningNoWerror = 0;
bit ShowInSystemHeader = 0;
+ bit Deferrable = 0;
Severity DefaultSeverity = defaultmapping;
DiagGroup Group;
string CategoryName = "";
@@ -98,6 +100,14 @@
bit ShowInSystemHeader = 0;
}
+class Deferrable {
+ bit Deferrable = 1;
+}
+
+class NonDeferrable {
+ bit Deferrable = 0;
+}
+
// FIXME: ExtWarn and Extension should also be SFINAEFailure by default.
class Error<string str> : Diagnostic<str, CLASS_ERROR, SEV_Error>, SFINAEFailure {
bit ShowInSystemHeader = 1;
Index: clang/test/SemaCUDA/deferred-oeverload.cu
===================================================================
--- /dev/null
+++ clang/test/SemaCUDA/deferred-oeverload.cu
@@ -0,0 +1,78 @@
+// RUN: %clang_cc1 -fcuda-is-device -fsyntax-only -verify=dev,com %s \
+// RUN: -std=c++11 -fgpu-defer-diag
+// RUN: %clang_cc1 -fsyntax-only -verify=host,com %s \
+// RUN: -std=c++11 -fgpu-defer-diag
+
+#include "Inputs/cuda.h"
+
+// When callee is called by a host function with integer arguments, there is an error for ambiguity.
+// It should be deferred since it involves wrong-sided candidates.
+__device__ void callee(int);
+__host__ void callee(float); // host-note {{candidate function}}
+__host__ void callee(double); // host-note {{candidate function}}
+
+// When callee2 is called by a device function without arguments, there is an error for 'no matching function'.
+// It should be deferred since it involves wrong-sided candidates.
+__host__ void callee2(); // dev-note{{candidate function not viable: call to __host__ function from __device__ function}}
+
+// When callee3 is called by a device function without arguments, there is an error for 'no matching function'.
+// It should be deferred since it involves wrong-sided candidates.
+__host__ void callee3(); // dev-note{{candidate function not viable: call to __host__ function from __device__ function}}
+__device__ void callee3(int); // dev-note{{candidate function not viable: requires 1 argument, but 0 were provided}}
+
+// When callee4 is called by a host or device function without arguments, there is an error for 'no matching function'.
+// It should be immediate since it involves no wrong-sided candidates (it is not a viable candiate due to signature).
+__host__ void callee4(int); // com-note 2{{candidate function not viable: requires 1 argument, but 0 were provided}}
+
+// When callee5 is called by a host function with integer arguments, there is an error for ambiguity.
+// It should be immediate since it involves no wrong-sided candidates.
+__host__ void callee5(float); // com-note {{candidate function}}
+__host__ void callee5(double); // com-note {{candidate function}}
+
+__host__ void hf() {
+ callee(1); // host-error {{call to 'callee' is ambiguous}}
+ callee2();
+ callee3();
+ callee4(); // com-error {{no matching function for call to 'callee4'}}
+ callee5(1); // com-error {{call to 'callee5' is ambiguous}}
+ undeclared_func(); // com-error {{use of undeclared identifier 'undeclared_func'}}
+}
+
+__device__ void df() {
+ callee(1);
+ callee2(); // dev-error {{no matching function for call to 'callee2'}}
+ callee3(); // dev-error {{no matching function for call to 'callee3'}}
+ callee4(); // com-error {{no matching function for call to 'callee4'}}
+}
+
+struct A { int x; typedef int isA; };
+struct B { int x; };
+
+// This function is invalid for A and B by SFINAE.
+// This fails to substitue for A but no diagnostic
+// should be emitted.
+template<typename T, typename T::foo* = nullptr>
+__host__ __device__ void sfinae(T t) { // com-note {{candidate template ignored: substitution failure [with T = B]}}
+ t.x = 1;
+}
+
+// This function is defined for A only by SFINAE.
+// Calling it with A should succeed, with B should fail.
+// The error should not be deferred since it happens in
+// file scope.
+
+template<typename T, typename T::isA* = nullptr>
+__host__ __device__ void sfinae(T t) { // com-note {{candidate template ignored: substitution failure [with T = B]}}
+ t.x = 1;
+}
+
+void test_sfinae() {
+ sfinae(A());
+ sfinae(B()); // com-error{{no matching function for call to 'sfinae'}}
+}
+
+// If a syntax error causes a function not declared, it cannot
+// be deferred.
+
+inline __host__ __device__ void bad_func() { // com-note {{to match this '{'}}
+// com-error {{expected '}'}}
Index: clang/lib/Sema/SemaType.cpp
===================================================================
--- clang/lib/Sema/SemaType.cpp
+++ clang/lib/Sema/SemaType.cpp
@@ -4131,7 +4131,8 @@
/// Creates a fix-it to insert a C-style nullability keyword at \p pointerLoc,
/// taking into account whitespace before and after.
-static void fixItNullability(Sema &S, DiagnosticBuilder &Diag,
+template <typename DiagBuilderT>
+static void fixItNullability(Sema &S, DiagBuilderT &Diag,
SourceLocation PointerLoc,
NullabilityKind Nullability) {
assert(PointerLoc.isValid());
Index: clang/lib/Sema/SemaTemplateVariadic.cpp
===================================================================
--- clang/lib/Sema/SemaTemplateVariadic.cpp
+++ clang/lib/Sema/SemaTemplateVariadic.cpp
@@ -368,8 +368,8 @@
Locations.push_back(Unexpanded[I].second);
}
- DiagnosticBuilder DB = Diag(Loc, diag::err_unexpanded_parameter_pack)
- << (int)UPPC << (int)Names.size();
+ auto DB = Diag(Loc, diag::err_unexpanded_parameter_pack)
+ << (int)UPPC << (int)Names.size();
for (size_t I = 0, E = std::min(Names.size(), (size_t)2); I != E; ++I)
DB << Names[I];
Index: clang/lib/Sema/SemaTemplateInstantiateDecl.cpp
===================================================================
--- clang/lib/Sema/SemaTemplateInstantiateDecl.cpp
+++ clang/lib/Sema/SemaTemplateInstantiateDecl.cpp
@@ -5979,7 +5979,7 @@
if (!Result) {
if (isa<UsingShadowDecl>(D)) {
// UsingShadowDecls can instantiate to nothing because of using hiding.
- } else if (Diags.hasUncompilableErrorOccurred()) {
+ } else if (hasUncompilableErrorOccurred()) {
// We've already complained about some ill-formed code, so most likely
// this declaration failed to instantiate. There's no point in
// complaining further, since this is normal in invalid code.
Index: clang/lib/Sema/SemaTemplateInstantiate.cpp
===================================================================
--- clang/lib/Sema/SemaTemplateInstantiate.cpp
+++ clang/lib/Sema/SemaTemplateInstantiate.cpp
@@ -237,7 +237,7 @@
// error have occurred. Any diagnostics we might have raised will not be
// visible, and we do not need to construct a correct AST.
if (SemaRef.Diags.hasFatalErrorOccurred() &&
- SemaRef.Diags.hasUncompilableErrorOccurred()) {
+ SemaRef.hasUncompilableErrorOccurred()) {
Invalid = true;
return;
}
Index: clang/lib/Sema/SemaStmtAsm.cpp
===================================================================
--- clang/lib/Sema/SemaStmtAsm.cpp
+++ clang/lib/Sema/SemaStmtAsm.cpp
@@ -448,9 +448,9 @@
unsigned Size = Context.getTypeSize(Ty);
if (!Context.getTargetInfo().validateInputSize(FeatureMap,
Literal->getString(), Size))
- return StmtResult(
- targetDiag(InputExpr->getBeginLoc(), diag::err_asm_invalid_input_size)
- << Info.getConstraintStr());
+ return targetDiag(InputExpr->getBeginLoc(),
+ diag::err_asm_invalid_input_size)
+ << Info.getConstraintStr();
}
// Check that the clobbers are valid.
Index: clang/lib/Sema/SemaStmt.cpp
===================================================================
--- clang/lib/Sema/SemaStmt.cpp
+++ clang/lib/Sema/SemaStmt.cpp
@@ -1249,10 +1249,10 @@
// Produce a nice diagnostic if multiple values aren't handled.
if (!UnhandledNames.empty()) {
- DiagnosticBuilder DB = Diag(CondExpr->getExprLoc(),
- TheDefaultStmt ? diag::warn_def_missing_case
+ auto DB = Diag(CondExpr->getExprLoc(), TheDefaultStmt
+ ? diag::warn_def_missing_case
: diag::warn_missing_case)
- << (int)UnhandledNames.size();
+ << (int)UnhandledNames.size();
for (size_t I = 0, E = std::min(UnhandledNames.size(), (size_t)3);
I != E; ++I)
Index: clang/lib/Sema/SemaSYCL.cpp
===================================================================
--- clang/lib/Sema/SemaSYCL.cpp
+++ clang/lib/Sema/SemaSYCL.cpp
@@ -17,19 +17,19 @@
// SYCL device specific diagnostics implementation
// -----------------------------------------------------------------------------
-Sema::DeviceDiagBuilder Sema::SYCLDiagIfDeviceCode(SourceLocation Loc,
- unsigned DiagID) {
+Sema::SemaDiagnosticBuilder Sema::SYCLDiagIfDeviceCode(SourceLocation Loc,
+ unsigned DiagID) {
assert(getLangOpts().SYCLIsDevice &&
"Should only be called during SYCL compilation");
FunctionDecl *FD = dyn_cast<FunctionDecl>(getCurLexicalContext());
- DeviceDiagBuilder::Kind DiagKind = [this, FD] {
+ SemaDiagnosticBuilder::Kind DiagKind = [this, FD] {
if (!FD)
- return DeviceDiagBuilder::K_Nop;
+ return SemaDiagnosticBuilder::K_Nop;
if (getEmissionStatus(FD) == Sema::FunctionEmissionStatus::Emitted)
- return DeviceDiagBuilder::K_ImmediateWithCallStack;
- return DeviceDiagBuilder::K_Deferred;
+ return SemaDiagnosticBuilder::K_ImmediateWithCallStack;
+ return SemaDiagnosticBuilder::K_Deferred;
}();
- return DeviceDiagBuilder(DiagKind, Loc, DiagID, FD, *this);
+ return SemaDiagnosticBuilder(DiagKind, Loc, DiagID, FD, *this);
}
bool Sema::checkSYCLDeviceFunction(SourceLocation Loc, FunctionDecl *Callee) {
@@ -42,8 +42,8 @@
if (isUnevaluatedContext() || isConstantEvaluated())
return true;
- DeviceDiagBuilder::Kind DiagKind = DeviceDiagBuilder::K_Nop;
+ SemaDiagnosticBuilder::Kind DiagKind = SemaDiagnosticBuilder::K_Nop;
- return DiagKind != DeviceDiagBuilder::K_Immediate &&
- DiagKind != DeviceDiagBuilder::K_ImmediateWithCallStack;
+ return DiagKind != SemaDiagnosticBuilder::K_Immediate &&
+ DiagKind != SemaDiagnosticBuilder::K_ImmediateWithCallStack;
}
Index: clang/lib/Sema/SemaOverload.cpp
===================================================================
--- clang/lib/Sema/SemaOverload.cpp
+++ clang/lib/Sema/SemaOverload.cpp
@@ -11528,9 +11528,18 @@
StringRef Opc, SourceLocation OpLoc,
llvm::function_ref<bool(OverloadCandidate &)> Filter) {
+ bool DeferHint = false;
+ if (S.getLangOpts().CUDA && S.getLangOpts().GPUDeferDiag) {
+ // Defer diagnostic for CUDA/HIP if there are wrong-sided candidates.
+ auto WrongSidedCands =
+ CompleteCandidates(S, OCD_AllCandidates, Args, OpLoc, [](auto &Cand) {
+ return Cand.FailureKind == ovl_fail_bad_target;
+ });
+ DeferHint = WrongSidedCands.size();
+ }
auto Cands = CompleteCandidates(S, OCD, Args, OpLoc, Filter);
- S.Diag(PD.first, PD.second);
+ S.Diag(PD.first, PD.second, DeferHint);
NoteCandidates(S, Args, Cands, Opc, OpLoc);
Index: clang/lib/Sema/SemaOpenMP.cpp
===================================================================
--- clang/lib/Sema/SemaOpenMP.cpp
+++ clang/lib/Sema/SemaOpenMP.cpp
@@ -1888,27 +1888,27 @@
};
} // anonymous namespace
-Sema::DeviceDiagBuilder Sema::diagIfOpenMPDeviceCode(SourceLocation Loc,
- unsigned DiagID) {
+Sema::SemaDiagnosticBuilder Sema::diagIfOpenMPDeviceCode(SourceLocation Loc,
+ unsigned DiagID) {
assert(LangOpts.OpenMP && LangOpts.OpenMPIsDevice &&
"Expected OpenMP device compilation.");
FunctionDecl *FD = getCurFunctionDecl();
- DeviceDiagBuilder::Kind Kind = DeviceDiagBuilder::K_Nop;
+ SemaDiagnosticBuilder::Kind Kind = SemaDiagnosticBuilder::K_Nop;
if (FD) {
FunctionEmissionStatus FES = getEmissionStatus(FD);
switch (FES) {
case FunctionEmissionStatus::Emitted:
- Kind = DeviceDiagBuilder::K_Immediate;
+ Kind = SemaDiagnosticBuilder::K_Immediate;
break;
case FunctionEmissionStatus::Unknown:
Kind = isOpenMPDeviceDelayedContext(*this)
- ? DeviceDiagBuilder::K_Deferred
- : DeviceDiagBuilder::K_Immediate;
+ ? SemaDiagnosticBuilder::K_Deferred
+ : SemaDiagnosticBuilder::K_Immediate;
break;
case FunctionEmissionStatus::TemplateDiscarded:
case FunctionEmissionStatus::OMPDiscarded:
- Kind = DeviceDiagBuilder::K_Nop;
+ Kind = SemaDiagnosticBuilder::K_Nop;
break;
case FunctionEmissionStatus::CUDADiscarded:
llvm_unreachable("CUDADiscarded unexpected in OpenMP device compilation");
@@ -1916,30 +1916,30 @@
}
}
- return DeviceDiagBuilder(Kind, Loc, DiagID, getCurFunctionDecl(), *this);
+ return SemaDiagnosticBuilder(Kind, Loc, DiagID, getCurFunctionDecl(), *this);
}
-Sema::DeviceDiagBuilder Sema::diagIfOpenMPHostCode(SourceLocation Loc,
- unsigned DiagID) {
+Sema::SemaDiagnosticBuilder Sema::diagIfOpenMPHostCode(SourceLocation Loc,
+ unsigned DiagID) {
assert(LangOpts.OpenMP && !LangOpts.OpenMPIsDevice &&
"Expected OpenMP host compilation.");
FunctionEmissionStatus FES = getEmissionStatus(getCurFunctionDecl());
- DeviceDiagBuilder::Kind Kind = DeviceDiagBuilder::K_Nop;
+ SemaDiagnosticBuilder::Kind Kind = SemaDiagnosticBuilder::K_Nop;
switch (FES) {
case FunctionEmissionStatus::Emitted:
- Kind = DeviceDiagBuilder::K_Immediate;
+ Kind = SemaDiagnosticBuilder::K_Immediate;
break;
case FunctionEmissionStatus::Unknown:
- Kind = DeviceDiagBuilder::K_Deferred;
+ Kind = SemaDiagnosticBuilder::K_Deferred;
break;
case FunctionEmissionStatus::TemplateDiscarded:
case FunctionEmissionStatus::OMPDiscarded:
case FunctionEmissionStatus::CUDADiscarded:
- Kind = DeviceDiagBuilder::K_Nop;
+ Kind = SemaDiagnosticBuilder::K_Nop;
break;
}
- return DeviceDiagBuilder(Kind, Loc, DiagID, getCurFunctionDecl(), *this);
+ return SemaDiagnosticBuilder(Kind, Loc, DiagID, getCurFunctionDecl(), *this);
}
static OpenMPDefaultmapClauseKind
Index: clang/lib/Sema/SemaExprObjC.cpp
===================================================================
--- clang/lib/Sema/SemaExprObjC.cpp
+++ clang/lib/Sema/SemaExprObjC.cpp
@@ -2445,8 +2445,8 @@
SourceManager &SM = S.SourceMgr;
edit::Commit ECommit(SM, S.LangOpts);
if (refactor(Msg,*S.NSAPIObj, ECommit)) {
- DiagnosticBuilder Builder = S.Diag(MsgLoc, DiagID)
- << Msg->getSelector() << Msg->getSourceRange();
+ auto Builder = S.Diag(MsgLoc, DiagID)
+ << Msg->getSelector() << Msg->getSourceRange();
// FIXME: Don't emit diagnostic at all if fixits are non-commitable.
if (!ECommit.isCommitable())
return;
@@ -3139,9 +3139,8 @@
if (ReceiverType->isObjCClassType() && !isImplicit &&
!(Receiver->isObjCSelfExpr() && getLangOpts().ObjCAutoRefCount)) {
{
- DiagnosticBuilder Builder =
- Diag(Receiver->getExprLoc(),
- diag::err_messaging_class_with_direct_method);
+ auto Builder = Diag(Receiver->getExprLoc(),
+ diag::err_messaging_class_with_direct_method);
if (Receiver->isObjCSelfExpr()) {
Builder.AddFixItHint(FixItHint::CreateReplacement(
RecRange, Method->getClassInterface()->getName()));
@@ -3153,7 +3152,7 @@
if (SuperLoc.isValid()) {
{
- DiagnosticBuilder Builder =
+ auto Builder =
Diag(SuperLoc, diag::err_messaging_super_with_direct_method);
if (ReceiverType->isObjCClassType()) {
Builder.AddFixItHint(FixItHint::CreateReplacement(
@@ -3736,15 +3735,11 @@
return LookupName(R, TUScope, false);
}
-static void addFixitForObjCARCConversion(Sema &S,
- DiagnosticBuilder &DiagB,
- Sema::CheckedConversionKind CCK,
- SourceLocation afterLParen,
- QualType castType,
- Expr *castExpr,
- Expr *realCast,
- const char *bridgeKeyword,
- const char *CFBridgeName) {
+template <typename DiagBuilderT>
+static void addFixitForObjCARCConversion(
+ Sema &S, DiagBuilderT &DiagB, Sema::CheckedConversionKind CCK,
+ SourceLocation afterLParen, QualType castType, Expr *castExpr,
+ Expr *realCast, const char *bridgeKeyword, const char *CFBridgeName) {
// We handle C-style and implicit casts here.
switch (CCK) {
case Sema::CCK_ImplicitConversion:
@@ -3921,9 +3916,9 @@
assert(CreateRule != ACC_bottom && "This cast should already be accepted.");
if (CreateRule != ACC_plusOne)
{
- DiagnosticBuilder DiagB =
- (CCK != Sema::CCK_OtherCast) ? S.Diag(noteLoc, diag::note_arc_bridge)
- : S.Diag(noteLoc, diag::note_arc_cstyle_bridge);
+ auto DiagB = (CCK != Sema::CCK_OtherCast)
+ ? S.Diag(noteLoc, diag::note_arc_bridge)
+ : S.Diag(noteLoc, diag::note_arc_cstyle_bridge);
addFixitForObjCARCConversion(S, DiagB, CCK, afterLParen,
castType, castExpr, realCast, "__bridge ",
@@ -3931,12 +3926,12 @@
}
if (CreateRule != ACC_plusZero)
{
- DiagnosticBuilder DiagB =
- (CCK == Sema::CCK_OtherCast && !br) ?
- S.Diag(noteLoc, diag::note_arc_cstyle_bridge_transfer) << castExprType :
- S.Diag(br ? castExpr->getExprLoc() : noteLoc,
- diag::note_arc_bridge_transfer)
- << castExprType << br;
+ auto DiagB = (CCK == Sema::CCK_OtherCast && !br)
+ ? S.Diag(noteLoc, diag::note_arc_cstyle_bridge_transfer)
+ << castExprType
+ : S.Diag(br ? castExpr->getExprLoc() : noteLoc,
+ diag::note_arc_bridge_transfer)
+ << castExprType << br;
addFixitForObjCARCConversion(S, DiagB, CCK, afterLParen,
castType, castExpr, realCast, "__bridge_transfer ",
@@ -3962,21 +3957,21 @@
assert(CreateRule != ACC_bottom && "This cast should already be accepted.");
if (CreateRule != ACC_plusOne)
{
- DiagnosticBuilder DiagB =
- (CCK != Sema::CCK_OtherCast) ? S.Diag(noteLoc, diag::note_arc_bridge)
- : S.Diag(noteLoc, diag::note_arc_cstyle_bridge);
+ auto DiagB = (CCK != Sema::CCK_OtherCast)
+ ? S.Diag(noteLoc, diag::note_arc_bridge)
+ : S.Diag(noteLoc, diag::note_arc_cstyle_bridge);
addFixitForObjCARCConversion(S, DiagB, CCK, afterLParen,
castType, castExpr, realCast, "__bridge ",
nullptr);
}
if (CreateRule != ACC_plusZero)
{
- DiagnosticBuilder DiagB =
- (CCK == Sema::CCK_OtherCast && !br) ?
- S.Diag(noteLoc, diag::note_arc_cstyle_bridge_retained) << castType :
- S.Diag(br ? castExpr->getExprLoc() : noteLoc,
- diag::note_arc_bridge_retained)
- << castType << br;
+ auto DiagB = (CCK == Sema::CCK_OtherCast && !br)
+ ? S.Diag(noteLoc, diag::note_arc_cstyle_bridge_retained)
+ << castType
+ : S.Diag(br ? castExpr->getExprLoc() : noteLoc,
+ diag::note_arc_bridge_retained)
+ << castType << br;
addFixitForObjCARCConversion(S, DiagB, CCK, afterLParen,
castType, castExpr, realCast, "__bridge_retained ",
Index: clang/lib/Sema/SemaDecl.cpp
===================================================================
--- clang/lib/Sema/SemaDecl.cpp
+++ clang/lib/Sema/SemaDecl.cpp
@@ -14542,11 +14542,11 @@
// If any errors have occurred, clear out any temporaries that may have
// been leftover. This ensures that these temporaries won't be picked up for
// deletion in some later function.
- if (getDiagnostics().hasUncompilableErrorOccurred() ||
+ if (hasUncompilableErrorOccurred() ||
getDiagnostics().getSuppressAllDiagnostics()) {
DiscardCleanupsInEvaluationContext();
}
- if (!getDiagnostics().hasUncompilableErrorOccurred() &&
+ if (!hasUncompilableErrorOccurred() &&
!isa<FunctionTemplateDecl>(dcl)) {
// Since the body is valid, issue any analysis-based warnings that are
// enabled.
@@ -14598,7 +14598,7 @@
// If any errors have occurred, clear out any temporaries that may have
// been leftover. This ensures that these temporaries won't be picked up for
// deletion in some later function.
- if (getDiagnostics().hasUncompilableErrorOccurred()) {
+ if (hasUncompilableErrorOccurred()) {
DiscardCleanupsInEvaluationContext();
}
Index: clang/lib/Sema/SemaCUDA.cpp
===================================================================
--- clang/lib/Sema/SemaCUDA.cpp
+++ clang/lib/Sema/SemaCUDA.cpp
@@ -639,58 +639,63 @@
}
}
-Sema::DeviceDiagBuilder Sema::CUDADiagIfDeviceCode(SourceLocation Loc,
- unsigned DiagID) {
+Sema::SemaDiagnosticBuilder Sema::CUDADiagIfDeviceCode(SourceLocation Loc,
+ unsigned DiagID) {
assert(getLangOpts().CUDA && "Should only be called during CUDA compilation");
- DeviceDiagBuilder::Kind DiagKind = [this] {
+ SemaDiagnosticBuilder::Kind DiagKind = [&] {
+ if (!isa<FunctionDecl>(CurContext))
+ return SemaDiagnosticBuilder::K_Immediate;
switch (CurrentCUDATarget()) {
case CFT_Global:
case CFT_Device:
- return DeviceDiagBuilder::K_Immediate;
+ return SemaDiagnosticBuilder::K_Immediate;
case CFT_HostDevice:
// An HD function counts as host code if we're compiling for host, and
// device code if we're compiling for device. Defer any errors in device
// mode until the function is known-emitted.
- if (getLangOpts().CUDAIsDevice) {
- return (getEmissionStatus(cast<FunctionDecl>(CurContext)) ==
- FunctionEmissionStatus::Emitted)
- ? DeviceDiagBuilder::K_ImmediateWithCallStack
- : DeviceDiagBuilder::K_Deferred;
- }
- return DeviceDiagBuilder::K_Nop;
-
+ if (!getLangOpts().CUDAIsDevice)
+ return SemaDiagnosticBuilder::K_Nop;
+ if (IsLastErrorImmediate && Diags.getDiagnosticIDs()->isBuiltinNote(DiagID))
+ return SemaDiagnosticBuilder::K_Immediate;
+ return (getEmissionStatus(cast<FunctionDecl>(CurContext)) ==
+ FunctionEmissionStatus::Emitted)
+ ? SemaDiagnosticBuilder::K_ImmediateWithCallStack
+ : SemaDiagnosticBuilder::K_Deferred;
default:
- return DeviceDiagBuilder::K_Nop;
+ return SemaDiagnosticBuilder::K_Nop;
}
}();
- return DeviceDiagBuilder(DiagKind, Loc, DiagID,
- dyn_cast<FunctionDecl>(CurContext), *this);
+ return SemaDiagnosticBuilder(DiagKind, Loc, DiagID,
+ dyn_cast<FunctionDecl>(CurContext), *this);
}
-Sema::DeviceDiagBuilder Sema::CUDADiagIfHostCode(SourceLocation Loc,
- unsigned DiagID) {
+Sema::SemaDiagnosticBuilder Sema::CUDADiagIfHostCode(SourceLocation Loc,
+ unsigned DiagID) {
assert(getLangOpts().CUDA && "Should only be called during CUDA compilation");
- DeviceDiagBuilder::Kind DiagKind = [this] {
+ SemaDiagnosticBuilder::Kind DiagKind = [&] {
+ if (!isa<FunctionDecl>(CurContext))
+ return SemaDiagnosticBuilder::K_Immediate;
switch (CurrentCUDATarget()) {
case CFT_Host:
- return DeviceDiagBuilder::K_Immediate;
+ return SemaDiagnosticBuilder::K_Immediate;
case CFT_HostDevice:
// An HD function counts as host code if we're compiling for host, and
// device code if we're compiling for device. Defer any errors in device
// mode until the function is known-emitted.
if (getLangOpts().CUDAIsDevice)
- return DeviceDiagBuilder::K_Nop;
-
+ return SemaDiagnosticBuilder::K_Nop;
+ if (IsLastErrorImmediate && Diags.getDiagnosticIDs()->isBuiltinNote(DiagID))
+ return SemaDiagnosticBuilder::K_Immediate;
return (getEmissionStatus(cast<FunctionDecl>(CurContext)) ==
FunctionEmissionStatus::Emitted)
- ? DeviceDiagBuilder::K_ImmediateWithCallStack
- : DeviceDiagBuilder::K_Deferred;
+ ? SemaDiagnosticBuilder::K_ImmediateWithCallStack
+ : SemaDiagnosticBuilder::K_Deferred;
default:
- return DeviceDiagBuilder::K_Nop;
+ return SemaDiagnosticBuilder::K_Nop;
}
}();
- return DeviceDiagBuilder(DiagKind, Loc, DiagID,
- dyn_cast<FunctionDecl>(CurContext), *this);
+ return SemaDiagnosticBuilder(DiagKind, Loc, DiagID,
+ dyn_cast<FunctionDecl>(CurContext), *this);
}
bool Sema::CheckCUDACall(SourceLocation Loc, FunctionDecl *Callee) {
@@ -711,8 +716,8 @@
// Otherwise, mark the call in our call graph so we can traverse it later.
bool CallerKnownEmitted =
getEmissionStatus(Caller) == FunctionEmissionStatus::Emitted;
- DeviceDiagBuilder::Kind DiagKind = [this, Caller, Callee,
- CallerKnownEmitted] {
+ SemaDiagnosticBuilder::Kind DiagKind = [this, Caller, Callee,
+ CallerKnownEmitted] {
switch (IdentifyCUDAPreference(Caller, Callee)) {
case CFP_Never:
case CFP_WrongSide:
@@ -720,14 +725,15 @@
// If we know the caller will be emitted, we know this wrong-side call
// will be emitted, so it's an immediate error. Otherwise, defer the
// error until we know the caller is emitted.
- return CallerKnownEmitted ? DeviceDiagBuilder::K_ImmediateWithCallStack
- : DeviceDiagBuilder::K_Deferred;
+ return CallerKnownEmitted
+ ? SemaDiagnosticBuilder::K_ImmediateWithCallStack
+ : SemaDiagnosticBuilder::K_Deferred;
default:
- return DeviceDiagBuilder::K_Nop;
+ return SemaDiagnosticBuilder::K_Nop;
}
}();
- if (DiagKind == DeviceDiagBuilder::K_Nop)
+ if (DiagKind == SemaDiagnosticBuilder::K_Nop)
return true;
// Avoid emitting this error twice for the same location. Using a hashtable
@@ -737,14 +743,14 @@
if (!LocsWithCUDACallDiags.insert({Caller, Loc}).second)
return true;
- DeviceDiagBuilder(DiagKind, Loc, diag::err_ref_bad_target, Caller, *this)
+ SemaDiagnosticBuilder(DiagKind, Loc, diag::err_ref_bad_target, Caller, *this)
<< IdentifyCUDATarget(Callee) << Callee << IdentifyCUDATarget(Caller);
if (!Callee->getBuiltinID())
- DeviceDiagBuilder(DiagKind, Callee->getLocation(), diag::note_previous_decl,
- Caller, *this)
+ SemaDiagnosticBuilder(DiagKind, Callee->getLocation(),
+ diag::note_previous_decl, Caller, *this)
<< Callee;
- return DiagKind != DeviceDiagBuilder::K_Immediate &&
- DiagKind != DeviceDiagBuilder::K_ImmediateWithCallStack;
+ return DiagKind != SemaDiagnosticBuilder::K_Immediate &&
+ DiagKind != SemaDiagnosticBuilder::K_ImmediateWithCallStack;
}
// Check the wrong-sided reference capture of lambda for CUDA/HIP.
@@ -781,14 +787,14 @@
bool ShouldCheck = CalleeIsDevice && CallerIsHost;
if (!ShouldCheck || !Capture.isReferenceCapture())
return;
- auto DiagKind = DeviceDiagBuilder::K_Deferred;
+ auto DiagKind = SemaDiagnosticBuilder::K_Deferred;
if (Capture.isVariableCapture()) {
- DeviceDiagBuilder(DiagKind, Capture.getLocation(),
- diag::err_capture_bad_target, Callee, *this)
+ SemaDiagnosticBuilder(DiagKind, Capture.getLocation(),
+ diag::err_capture_bad_target, Callee, *this)
<< Capture.getVariable();
} else if (Capture.isThisCapture()) {
- DeviceDiagBuilder(DiagKind, Capture.getLocation(),
- diag::err_capture_bad_target_this_ptr, Callee, *this);
+ SemaDiagnosticBuilder(DiagKind, Capture.getLocation(),
+ diag::err_capture_bad_target_this_ptr, Callee, *this);
}
return;
}
Index: clang/lib/Sema/SemaAttr.cpp
===================================================================
--- clang/lib/Sema/SemaAttr.cpp
+++ clang/lib/Sema/SemaAttr.cpp
@@ -380,8 +380,8 @@
// The user might have already reset the alignment, so suggest replacing
// the reset with a pop.
if (IsInnermost && PackStack.CurrentValue == PackStack.DefaultValue) {
- DiagnosticBuilder DB = Diag(PackStack.CurrentPragmaLocation,
- diag::note_pragma_pack_pop_instead_reset);
+ auto DB = Diag(PackStack.CurrentPragmaLocation,
+ diag::note_pragma_pack_pop_instead_reset);
SourceLocation FixItLoc = Lexer::findLocationAfterToken(
PackStack.CurrentPragmaLocation, tok::l_paren, SourceMgr, LangOpts,
/*SkipTrailing=*/false);
Index: clang/lib/Sema/Sema.cpp
===================================================================
--- clang/lib/Sema/Sema.cpp
+++ clang/lib/Sema/Sema.cpp
@@ -1435,11 +1435,24 @@
}
Sema::SemaDiagnosticBuilder
-Sema::Diag(SourceLocation Loc, const PartialDiagnostic& PD) {
- SemaDiagnosticBuilder Builder(Diag(Loc, PD.getDiagID()));
- PD.Emit(Builder);
+Sema::Diag(SourceLocation Loc, const PartialDiagnostic &PD, bool DeferHint) {
+ return Diag(Loc, PD.getDiagID(), DeferHint) << PD;
+}
- return Builder;
+bool Sema::hasUncompilableErrorOccurred() const {
+ if (getDiagnostics().hasUncompilableErrorOccurred())
+ return true;
+ auto *FD = dyn_cast<FunctionDecl>(CurContext);
+ if (!FD)
+ return false;
+ auto Loc = DeviceDeferredDiags.find(FD);
+ if (Loc == DeviceDeferredDiags.end())
+ return false;
+ for (auto PDAt : Loc->second) {
+ if (DiagnosticIDs::isDefaultMappingAsError(PDAt.second.getDiagID()))
+ return true;
+ }
+ return false;
}
// Print notes showing how we can reach FD starting from an a priori
@@ -1652,9 +1665,9 @@
// until we discover that the function is known-emitted, at which point we take
// it out of this map and emit the diagnostic.
-Sema::DeviceDiagBuilder::DeviceDiagBuilder(Kind K, SourceLocation Loc,
- unsigned DiagID, FunctionDecl *Fn,
- Sema &S)
+Sema::SemaDiagnosticBuilder::SemaDiagnosticBuilder(Kind K, SourceLocation Loc,
+ unsigned DiagID,
+ FunctionDecl *Fn, Sema &S)
: S(S), Loc(Loc), DiagID(DiagID), Fn(Fn),
ShowCallStack(K == K_ImmediateWithCallStack || K == K_Deferred) {
switch (K) {
@@ -1662,7 +1675,8 @@
break;
case K_Immediate:
case K_ImmediateWithCallStack:
- ImmediateDiag.emplace(S.Diag(Loc, DiagID));
+ ImmediateDiag.emplace(
+ ImmediateDiagBuilder(S.Diags.Report(Loc, DiagID), S, DiagID));
break;
case K_Deferred:
assert(Fn && "Must have a function to attach the deferred diag to.");
@@ -1673,7 +1687,7 @@
}
}
-Sema::DeviceDiagBuilder::DeviceDiagBuilder(DeviceDiagBuilder &&D)
+Sema::SemaDiagnosticBuilder::SemaDiagnosticBuilder(SemaDiagnosticBuilder &&D)
: S(D.S), Loc(D.Loc), DiagID(D.DiagID), Fn(D.Fn),
ShowCallStack(D.ShowCallStack), ImmediateDiag(D.ImmediateDiag),
PartialDiagId(D.PartialDiagId) {
@@ -1683,7 +1697,7 @@
D.PartialDiagId.reset();
}
-Sema::DeviceDiagBuilder::~DeviceDiagBuilder() {
+Sema::SemaDiagnosticBuilder::~SemaDiagnosticBuilder() {
if (ImmediateDiag) {
// Emit our diagnostic and, if it was a warning or error, output a callstack
// if Fn isn't a priori known-emitted.
@@ -1698,7 +1712,8 @@
}
}
-Sema::DeviceDiagBuilder Sema::targetDiag(SourceLocation Loc, unsigned DiagID) {
+Sema::SemaDiagnosticBuilder Sema::targetDiag(SourceLocation Loc,
+ unsigned DiagID) {
if (LangOpts.OpenMP)
return LangOpts.OpenMPIsDevice ? diagIfOpenMPDeviceCode(Loc, DiagID)
: diagIfOpenMPHostCode(Loc, DiagID);
@@ -1709,8 +1724,32 @@
if (getLangOpts().SYCLIsDevice)
return SYCLDiagIfDeviceCode(Loc, DiagID);
- return DeviceDiagBuilder(DeviceDiagBuilder::K_Immediate, Loc, DiagID,
- getCurFunctionDecl(), *this);
+ return SemaDiagnosticBuilder(SemaDiagnosticBuilder::K_Immediate, Loc, DiagID,
+ getCurFunctionDecl(), *this);
+}
+
+Sema::SemaDiagnosticBuilder Sema::Diag(SourceLocation Loc, unsigned DiagID,
+ bool DeferHint) {
+ bool IsError = Diags.getDiagnosticIDs()->isDefaultMappingAsError(DiagID);
+ bool ShouldDefer = getLangOpts().CUDA && LangOpts.GPUDeferDiag &&
+ DiagnosticIDs::isDeferrable(DiagID) &&
+ (DeferHint || !IsError);
+ auto SetIsLastErrorImmediate = [&](bool Flag) {
+ if (IsError)
+ IsLastErrorImmediate = Flag;
+ };
+ if (!ShouldDefer) {
+ SetIsLastErrorImmediate(true);
+ return SemaDiagnosticBuilder(SemaDiagnosticBuilder::K_Immediate, Loc,
+ DiagID, getCurFunctionDecl(), *this);
+ }
+
+ SemaDiagnosticBuilder DB =
+ getLangOpts().CUDAIsDevice
+ ? CUDADiagIfDeviceCode(Loc, DiagID)
+ : CUDADiagIfHostCode(Loc, DiagID);
+ SetIsLastErrorImmediate(DB.isImmediate());
+ return DB;
}
void Sema::checkDeviceDecl(const ValueDecl *D, SourceLocation Loc) {
Index: clang/lib/Sema/AnalysisBasedWarnings.cpp
===================================================================
--- clang/lib/Sema/AnalysisBasedWarnings.cpp
+++ clang/lib/Sema/AnalysisBasedWarnings.cpp
@@ -2096,7 +2096,7 @@
if (cast<DeclContext>(D)->isDependentContext())
return;
- if (Diags.hasUncompilableErrorOccurred()) {
+ if (S.hasUncompilableErrorOccurred()) {
// Flush out any possibly unreachable diagnostics.
flushDiagnostics(S, fscope);
return;
Index: clang/lib/Frontend/CompilerInvocation.cpp
===================================================================
--- clang/lib/Frontend/CompilerInvocation.cpp
+++ clang/lib/Frontend/CompilerInvocation.cpp
@@ -2628,6 +2628,9 @@
if (Args.hasArg(OPT_fno_cuda_host_device_constexpr))
Opts.CUDAHostDeviceConstexpr = 0;
+ if (Args.hasArg(OPT_fgpu_defer_diag))
+ Opts.GPUDeferDiag = 1;
+
if (Opts.CUDAIsDevice && Args.hasArg(OPT_fcuda_approx_transcendentals))
Opts.CUDADeviceApproxTranscendentals = 1;
Index: clang/lib/Driver/ToolChains/HIP.cpp
===================================================================
--- clang/lib/Driver/ToolChains/HIP.cpp
+++ clang/lib/Driver/ToolChains/HIP.cpp
@@ -266,6 +266,10 @@
options::OPT_fno_gpu_allow_device_init, false))
CC1Args.push_back("-fgpu-allow-device-init");
+ if (DriverArgs.hasFlag(options::OPT_fgpu_defer_diag,
+ options::OPT_fno_gpu_defer_diag, false))
+ CC1Args.push_back("-fgpu-defer-diag");
+
CC1Args.push_back("-fcuda-allow-variadic-functions");
// Default to "hidden" visibility, as object level linking will not be
Index: clang/lib/Driver/ToolChains/Cuda.cpp
===================================================================
--- clang/lib/Driver/ToolChains/Cuda.cpp
+++ clang/lib/Driver/ToolChains/Cuda.cpp
@@ -634,6 +634,10 @@
if (DriverArgs.hasFlag(options::OPT_fgpu_rdc, options::OPT_fno_gpu_rdc,
false))
CC1Args.push_back("-fgpu-rdc");
+
+ if (DriverArgs.hasFlag(options::OPT_fgpu_defer_diag,
+ options::OPT_fno_gpu_defer_diag, false))
+ CC1Args.push_back("-fgpu-defer-diag");
}
if (DriverArgs.hasArg(options::OPT_nogpulib))
Index: clang/lib/Basic/DiagnosticIDs.cpp
===================================================================
--- clang/lib/Basic/DiagnosticIDs.cpp
+++ clang/lib/Basic/DiagnosticIDs.cpp
@@ -42,6 +42,7 @@
unsigned SFINAE : 2;
unsigned WarnNoWerror : 1;
unsigned WarnShowInSystemHeader : 1;
+ unsigned Deferrable : 1;
unsigned Category : 6;
uint16_t OptionGroupIndex;
@@ -96,12 +97,10 @@
static const StaticDiagInfoRec StaticDiagInfo[] = {
#define DIAG(ENUM, CLASS, DEFAULT_SEVERITY, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
- { \
- diag::ENUM, DEFAULT_SEVERITY, CLASS, DiagnosticIDs::SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY, GROUP, STR_SIZE(DESC, uint16_t), DESC \
- } \
- ,
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
+ {diag::ENUM, DEFAULT_SEVERITY, CLASS, DiagnosticIDs::SFINAE, \
+ NOWERROR, SHOWINSYSHEADER, DEFERRABLE, CATEGORY, \
+ GROUP, STR_SIZE(DESC, uint16_t), DESC},
#include "clang/Basic/DiagnosticCommonKinds.inc"
#include "clang/Basic/DiagnosticDriverKinds.inc"
#include "clang/Basic/DiagnosticFrontendKinds.inc"
@@ -253,6 +252,12 @@
return SFINAE_Report;
}
+bool DiagnosticIDs::isDeferrable(unsigned DiagID) {
+ if (const StaticDiagInfoRec *Info = GetDiagInfo(DiagID))
+ return Info->Deferrable;
+ return false;
+}
+
/// getBuiltinDiagClass - Return the class field of the diagnostic.
///
static unsigned getBuiltinDiagClass(unsigned DiagID) {
Index: clang/include/clang/Sema/Sema.h
===================================================================
--- clang/include/clang/Sema/Sema.h
+++ clang/include/clang/Sema/Sema.h
@@ -1462,28 +1462,31 @@
/// template instantiation stacks.
///
/// This class provides a wrapper around the basic DiagnosticBuilder
- /// class that emits diagnostics. SemaDiagnosticBuilder is
+ /// class that emits diagnostics. ImmediateDiagBuilder is
/// responsible for emitting the diagnostic (as DiagnosticBuilder
/// does) and, if the diagnostic comes from inside a template
/// instantiation, printing the template instantiation stack as
/// well.
- class SemaDiagnosticBuilder : public DiagnosticBuilder {
+ class ImmediateDiagBuilder : public DiagnosticBuilder {
Sema &SemaRef;
unsigned DiagID;
public:
- SemaDiagnosticBuilder(DiagnosticBuilder &DB, Sema &SemaRef, unsigned DiagID)
- : DiagnosticBuilder(DB), SemaRef(SemaRef), DiagID(DiagID) { }
+ const static bool IsDiagBuilder = false;
+ ImmediateDiagBuilder(DiagnosticBuilder &DB, Sema &SemaRef, unsigned DiagID)
+ : DiagnosticBuilder(DB), SemaRef(SemaRef), DiagID(DiagID) {}
+ ImmediateDiagBuilder(DiagnosticBuilder &&DB, Sema &SemaRef, unsigned DiagID)
+ : DiagnosticBuilder(DB), SemaRef(SemaRef), DiagID(DiagID) {}
// This is a cunning lie. DiagnosticBuilder actually performs move
// construction in its copy constructor (but due to varied uses, it's not
// possible to conveniently express this as actual move construction). So
// the default copy ctor here is fine, because the base class disables the
- // source anyway, so the user-defined ~SemaDiagnosticBuilder is a safe no-op
+ // source anyway, so the user-defined ~ImmediateDiagBuilder is a safe no-op
// in that case anwyay.
- SemaDiagnosticBuilder(const SemaDiagnosticBuilder&) = default;
+ ImmediateDiagBuilder(const ImmediateDiagBuilder &) = default;
- ~SemaDiagnosticBuilder() {
+ ~ImmediateDiagBuilder() {
// If we aren't active, there is nothing to do.
if (!isActive()) return;
@@ -1504,28 +1507,138 @@
}
/// Teach operator<< to produce an object of the correct type.
- template<typename T>
- friend const SemaDiagnosticBuilder &operator<<(
- const SemaDiagnosticBuilder &Diag, const T &Value) {
+ template <typename T>
+ friend const ImmediateDiagBuilder &
+ operator<<(const ImmediateDiagBuilder &Diag, const T &Value) {
const DiagnosticBuilder &BaseDiag = Diag;
BaseDiag << Value;
return Diag;
}
- const static bool IsDiagBuilder = false;
};
+ /// A generic diagnostic builder for errors which may or may not be deferred.
+ ///
+ /// In CUDA, there exist constructs (e.g. variable-length arrays, try/catch)
+ /// which are not allowed to appear inside __device__ functions and are
+ /// allowed to appear in __host__ __device__ functions only if the host+device
+ /// function is never codegen'ed.
+ ///
+ /// To handle this, we use the notion of "deferred diagnostics", where we
+ /// attach a diagnostic to a FunctionDecl that's emitted iff it's codegen'ed.
+ ///
+ /// This class lets you emit either a regular diagnostic, a deferred
+ /// diagnostic, or no diagnostic at all, according to an argument you pass to
+ /// its constructor, thus simplifying the process of creating these "maybe
+ /// deferred" diagnostics.
+ class SemaDiagnosticBuilder {
+ public:
+ enum Kind {
+ /// Emit no diagnostics.
+ K_Nop,
+ /// Emit the diagnostic immediately (i.e., behave like Sema::Diag()).
+ K_Immediate,
+ /// Emit the diagnostic immediately, and, if it's a warning or error, also
+ /// emit a call stack showing how this function can be reached by an a
+ /// priori known-emitted function.
+ K_ImmediateWithCallStack,
+ /// Create a deferred diagnostic, which is emitted only if the function
+ /// it's attached to is codegen'ed. Also emit a call stack as with
+ /// K_ImmediateWithCallStack.
+ K_Deferred
+ };
+
+ SemaDiagnosticBuilder(Kind K, SourceLocation Loc, unsigned DiagID,
+ FunctionDecl *Fn, Sema &S);
+ SemaDiagnosticBuilder(SemaDiagnosticBuilder &&D);
+ SemaDiagnosticBuilder(const SemaDiagnosticBuilder &) = default;
+ ~SemaDiagnosticBuilder();
+
+ bool isImmediate() const { return ImmediateDiag.hasValue(); }
+
+ /// Convertible to bool: True if we immediately emitted an error, false if
+ /// we didn't emit an error or we created a deferred error.
+ ///
+ /// Example usage:
+ ///
+ /// if (SemaDiagnosticBuilder(...) << foo << bar)
+ /// return ExprError();
+ ///
+ /// But see CUDADiagIfDeviceCode() and CUDADiagIfHostCode() -- you probably
+ /// want to use these instead of creating a SemaDiagnosticBuilder yourself.
+ operator bool() const { return isImmediate(); }
+
+ template <typename T>
+ friend const SemaDiagnosticBuilder &
+ operator<<(const SemaDiagnosticBuilder &Diag, const T &Value) {
+ if (Diag.ImmediateDiag.hasValue())
+ *Diag.ImmediateDiag << Value;
+ else if (Diag.PartialDiagId.hasValue())
+ Diag.S.DeviceDeferredDiags[Diag.Fn][*Diag.PartialDiagId].second
+ << Value;
+ return Diag;
+ }
+
+ friend const SemaDiagnosticBuilder &
+ operator<<(const SemaDiagnosticBuilder &Diag, const PartialDiagnostic &PD) {
+ if (Diag.ImmediateDiag.hasValue())
+ PD.Emit(*Diag.ImmediateDiag);
+ else if (Diag.PartialDiagId.hasValue())
+ Diag.S.DeviceDeferredDiags[Diag.Fn][*Diag.PartialDiagId].second = PD;
+ return Diag;
+ }
+
+ void AddFixItHint(const FixItHint &Hint) const {
+ if (ImmediateDiag.hasValue())
+ ImmediateDiag->AddFixItHint(Hint);
+ else if (PartialDiagId.hasValue())
+ S.DeviceDeferredDiags[Fn][*PartialDiagId].second.AddFixItHint(Hint);
+ }
+
+ friend ExprResult ExprError(const SemaDiagnosticBuilder &) {
+ return ExprError();
+ }
+ friend StmtResult StmtError(const SemaDiagnosticBuilder &) {
+ return StmtError();
+ }
+ operator ExprResult() const { return ExprError(); }
+ operator StmtResult() const { return StmtError(); }
+ operator TypeResult() const { return TypeError(); }
+ operator DeclResult() const { return DeclResult(true); }
+ operator MemInitResult() const { return MemInitResult(true); }
+
+ private:
+ Sema &S;
+ SourceLocation Loc;
+ unsigned DiagID;
+ FunctionDecl *Fn;
+ bool ShowCallStack;
+
+ // Invariant: At most one of these Optionals has a value.
+ // FIXME: Switch these to a Variant once that exists.
+ llvm::Optional<ImmediateDiagBuilder> ImmediateDiag;
+ llvm::Optional<unsigned> PartialDiagId;
+ };
+ using DiagBuilderT = SemaDiagnosticBuilder;
+
+ /// Is the last error level diagnostic immediate. This is used to determined
+ /// whether the next info diagnostic should be immediate.
+ bool IsLastErrorImmediate = true;
+
/// Emit a diagnostic.
- SemaDiagnosticBuilder Diag(SourceLocation Loc, unsigned DiagID) {
- DiagnosticBuilder DB = Diags.Report(Loc, DiagID);
- return SemaDiagnosticBuilder(DB, *this, DiagID);
- }
+ SemaDiagnosticBuilder Diag(SourceLocation Loc, unsigned DiagID,
+ bool DeferHint = false);
/// Emit a partial diagnostic.
- SemaDiagnosticBuilder Diag(SourceLocation Loc, const PartialDiagnostic& PD);
+ SemaDiagnosticBuilder Diag(SourceLocation Loc, const PartialDiagnostic &PD,
+ bool DeferHint = false);
/// Build a partial diagnostic.
PartialDiagnostic PDiag(unsigned DiagID = 0); // in SemaInternal.h
+ /// Whether uncompilable error has occurred. This includes error happens
+ /// in deferred diagnostics.
+ bool hasUncompilableErrorOccurred() const;
+
bool findMacroSpelling(SourceLocation &loc, StringRef name);
/// Get a string to suggest for zero-initialization of a type.
@@ -11659,84 +11772,11 @@
/* Caller = */ FunctionDeclAndLoc>
DeviceKnownEmittedFns;
- /// Diagnostic builder for CUDA/OpenMP devices errors which may or may not be
- /// deferred.
- ///
- /// In CUDA, there exist constructs (e.g. variable-length arrays, try/catch)
- /// which are not allowed to appear inside __device__ functions and are
- /// allowed to appear in __host__ __device__ functions only if the host+device
- /// function is never codegen'ed.
- ///
- /// To handle this, we use the notion of "deferred diagnostics", where we
- /// attach a diagnostic to a FunctionDecl that's emitted iff it's codegen'ed.
- ///
- /// This class lets you emit either a regular diagnostic, a deferred
- /// diagnostic, or no diagnostic at all, according to an argument you pass to
- /// its constructor, thus simplifying the process of creating these "maybe
- /// deferred" diagnostics.
- class DeviceDiagBuilder {
- public:
- enum Kind {
- /// Emit no diagnostics.
- K_Nop,
- /// Emit the diagnostic immediately (i.e., behave like Sema::Diag()).
- K_Immediate,
- /// Emit the diagnostic immediately, and, if it's a warning or error, also
- /// emit a call stack showing how this function can be reached by an a
- /// priori known-emitted function.
- K_ImmediateWithCallStack,
- /// Create a deferred diagnostic, which is emitted only if the function
- /// it's attached to is codegen'ed. Also emit a call stack as with
- /// K_ImmediateWithCallStack.
- K_Deferred
- };
-
- DeviceDiagBuilder(Kind K, SourceLocation Loc, unsigned DiagID,
- FunctionDecl *Fn, Sema &S);
- DeviceDiagBuilder(DeviceDiagBuilder &&D);
- DeviceDiagBuilder(const DeviceDiagBuilder &) = default;
- ~DeviceDiagBuilder();
-
- /// Convertible to bool: True if we immediately emitted an error, false if
- /// we didn't emit an error or we created a deferred error.
- ///
- /// Example usage:
- ///
- /// if (DeviceDiagBuilder(...) << foo << bar)
- /// return ExprError();
- ///
- /// But see CUDADiagIfDeviceCode() and CUDADiagIfHostCode() -- you probably
- /// want to use these instead of creating a DeviceDiagBuilder yourself.
- operator bool() const { return ImmediateDiag.hasValue(); }
-
- template <typename T>
- friend const DeviceDiagBuilder &operator<<(const DeviceDiagBuilder &Diag,
- const T &Value) {
- if (Diag.ImmediateDiag.hasValue())
- *Diag.ImmediateDiag << Value;
- else if (Diag.PartialDiagId.hasValue())
- Diag.S.DeviceDeferredDiags[Diag.Fn][*Diag.PartialDiagId].second
- << Value;
- return Diag;
- }
-
- private:
- Sema &S;
- SourceLocation Loc;
- unsigned DiagID;
- FunctionDecl *Fn;
- bool ShowCallStack;
-
- // Invariant: At most one of these Optionals has a value.
- // FIXME: Switch these to a Variant once that exists.
- llvm::Optional<SemaDiagnosticBuilder> ImmediateDiag;
- llvm::Optional<unsigned> PartialDiagId;
- };
-
- /// Creates a DeviceDiagBuilder that emits the diagnostic if the current context
- /// is "used as device code".
+ /// Creates a SemaDiagnosticBuilder that emits the diagnostic if the current
+ /// context is "used as device code".
///
- /// - If CurContext is a __host__ function, does not emit any diagnostics.
+ /// - If CurContext is a __host__ function, does not emit any diagnostics
+ /// unless \p EmitOnBothSides is true.
/// - If CurContext is a __device__ or __global__ function, emits the
/// diagnostics immediately.
/// - If CurContext is a __host__ __device__ function and we are compiling for
@@ -11749,15 +11789,16 @@
/// if (CUDADiagIfDeviceCode(Loc, diag::err_cuda_vla) << CurrentCUDATarget())
/// return ExprError();
/// // Otherwise, continue parsing as normal.
- DeviceDiagBuilder CUDADiagIfDeviceCode(SourceLocation Loc, unsigned DiagID);
+ SemaDiagnosticBuilder CUDADiagIfDeviceCode(SourceLocation Loc,
+ unsigned DiagID);
- /// Creates a DeviceDiagBuilder that emits the diagnostic if the current context
- /// is "used as host code".
+ /// Creates a SemaDiagnosticBuilder that emits the diagnostic if the current
+ /// context is "used as host code".
///
/// Same as CUDADiagIfDeviceCode, with "host" and "device" switched.
- DeviceDiagBuilder CUDADiagIfHostCode(SourceLocation Loc, unsigned DiagID);
+ SemaDiagnosticBuilder CUDADiagIfHostCode(SourceLocation Loc, unsigned DiagID);
- /// Creates a DeviceDiagBuilder that emits the diagnostic if the current
+ /// Creates a SemaDiagnosticBuilder that emits the diagnostic if the current
/// context is "used as device code".
///
/// - If CurContext is a `declare target` function or it is known that the
@@ -11772,9 +11813,10 @@
/// if (diagIfOpenMPDeviceCode(Loc, diag::err_vla_unsupported))
/// return ExprError();
/// // Otherwise, continue parsing as normal.
- DeviceDiagBuilder diagIfOpenMPDeviceCode(SourceLocation Loc, unsigned DiagID);
+ SemaDiagnosticBuilder diagIfOpenMPDeviceCode(SourceLocation Loc,
+ unsigned DiagID);
- /// Creates a DeviceDiagBuilder that emits the diagnostic if the current
+ /// Creates a SemaDiagnosticBuilder that emits the diagnostic if the current
/// context is "used as host code".
///
/// - If CurContext is a `declare target` function or it is known that the
@@ -11787,9 +11829,14 @@
/// if (diagIfOpenMPHostode(Loc, diag::err_vla_unsupported))
/// return ExprError();
/// // Otherwise, continue parsing as normal.
- DeviceDiagBuilder diagIfOpenMPHostCode(SourceLocation Loc, unsigned DiagID);
+ SemaDiagnosticBuilder diagIfOpenMPHostCode(SourceLocation Loc,
+ unsigned DiagID);
- DeviceDiagBuilder targetDiag(SourceLocation Loc, unsigned DiagID);
+ SemaDiagnosticBuilder targetDiag(SourceLocation Loc, unsigned DiagID);
+ SemaDiagnosticBuilder targetDiag(SourceLocation Loc,
+ const PartialDiagnostic &PD) {
+ return targetDiag(Loc, PD.getDiagID()) << PD;
+ }
/// Check if the expression is allowed to be used in expressions for the
/// offloading devices.
@@ -12561,7 +12608,7 @@
ConstructorDestructor,
BuiltinFunction
};
- /// Creates a DeviceDiagBuilder that emits the diagnostic if the current
+ /// Creates a SemaDiagnosticBuilder that emits the diagnostic if the current
/// context is "used as device code".
///
/// - If CurLexicalContext is a kernel function or it is known that the
@@ -12579,7 +12626,8 @@
/// if (!S.Context.getTargetInfo().hasFloat128Type() &&
/// S.getLangOpts().SYCLIsDevice)
/// SYCLDiagIfDeviceCode(Loc, diag::err_type_unsupported) << "__float128";
- DeviceDiagBuilder SYCLDiagIfDeviceCode(SourceLocation Loc, unsigned DiagID);
+ SemaDiagnosticBuilder SYCLDiagIfDeviceCode(SourceLocation Loc,
+ unsigned DiagID);
/// Check whether we're allowed to call Callee from the current context.
///
Index: clang/include/clang/Driver/Options.td
===================================================================
--- clang/include/clang/Driver/Options.td
+++ clang/include/clang/Driver/Options.td
@@ -662,6 +662,9 @@
"Use", "Don't use", " new kernel launching API for HIP">;
defm gpu_allow_device_init : OptInFFlag<"gpu-allow-device-init",
"Allow", "Don't allow", " device side init function in HIP">;
+defm gpu_defer_diag : OptInFFlag<"gpu-defer-diag",
+ "Defer", "Don't defer", " all semantic diagnostics in host device functions"
+ " for CUDA/HIP">;
def gpu_max_threads_per_block_EQ : Joined<["--"], "gpu-max-threads-per-block=">,
Flags<[CC1Option]>,
HelpText<"Default max threads per block for kernel launch bounds for HIP">;
Index: clang/include/clang/Basic/LangOptions.def
===================================================================
--- clang/include/clang/Basic/LangOptions.def
+++ clang/include/clang/Basic/LangOptions.def
@@ -241,6 +241,7 @@
LANGOPT(GPURelocatableDeviceCode, 1, 0, "generate relocatable device code")
LANGOPT(GPUAllowDeviceInit, 1, 0, "allowing device side global init functions for HIP")
LANGOPT(GPUMaxThreadsPerBlock, 32, 256, "default max threads per block for kernel launch bounds for HIP")
+LANGOPT(GPUDeferDiag, 1, 0, "defer all semantic diagnostic messages in host device functions for CUDA/HIP")
LANGOPT(SYCL , 1, 0, "SYCL")
LANGOPT(SYCLIsDevice , 1, 0, "Generate code for SYCL device")
Index: clang/include/clang/Basic/DiagnosticSerialization.h
===================================================================
--- clang/include/clang/Basic/DiagnosticSerialization.h
+++ clang/include/clang/Basic/DiagnosticSerialization.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define SERIALIZATIONSTART
#include "clang/Basic/DiagnosticSerializationKinds.inc"
Index: clang/include/clang/Basic/DiagnosticSemaKinds.td
===================================================================
--- clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -4050,6 +4050,8 @@
"static and non-static member functions with the same parameter types "
"cannot be overloaded">;
+let Deferrable = 1 in {
+
def err_ovl_no_viable_function_in_call : Error<
"no matching function for call to %0">;
def err_ovl_no_viable_member_function_in_call : Error<
@@ -4363,6 +4365,8 @@
def err_addr_ovl_no_qualifier : Error<
"cannot form member pointer of type %0 without '&' and class name">;
+} // let Deferrable
+
// C++11 Literal Operators
def err_ovl_no_viable_literal_operator : Error<
"no matching literal operator for call to %0"
Index: clang/include/clang/Basic/DiagnosticSema.h
===================================================================
--- clang/include/clang/Basic/DiagnosticSema.h
+++ clang/include/clang/Basic/DiagnosticSema.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define SEMASTART
#include "clang/Basic/DiagnosticSemaKinds.inc"
Index: clang/include/clang/Basic/DiagnosticRefactoring.h
===================================================================
--- clang/include/clang/Basic/DiagnosticRefactoring.h
+++ clang/include/clang/Basic/DiagnosticRefactoring.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define REFACTORINGSTART
#include "clang/Basic/DiagnosticRefactoringKinds.inc"
Index: clang/include/clang/Basic/DiagnosticParse.h
===================================================================
--- clang/include/clang/Basic/DiagnosticParse.h
+++ clang/include/clang/Basic/DiagnosticParse.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define PARSESTART
#include "clang/Basic/DiagnosticParseKinds.inc"
Index: clang/include/clang/Basic/DiagnosticLex.h
===================================================================
--- clang/include/clang/Basic/DiagnosticLex.h
+++ clang/include/clang/Basic/DiagnosticLex.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define LEXSTART
#include "clang/Basic/DiagnosticLexKinds.inc"
Index: clang/include/clang/Basic/DiagnosticIDs.h
===================================================================
--- clang/include/clang/Basic/DiagnosticIDs.h
+++ clang/include/clang/Basic/DiagnosticIDs.h
@@ -64,8 +64,9 @@
// Get typedefs for common diagnostics.
enum {
-#define DIAG(ENUM,FLAGS,DEFAULT_MAPPING,DESC,GROUP,\
- SFINAE,CATEGORY,NOWERROR,SHOWINSYSHEADER) ENUM,
+#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, CATEGORY, \
+ NOWERROR, SHOWINSYSHEADER, DEFFERABLE) \
+ ENUM,
#define COMMONSTART
#include "clang/Basic/DiagnosticCommonKinds.inc"
NUM_BUILTIN_COMMON_DIAGNOSTICS
@@ -280,6 +281,13 @@
/// are not SFINAE errors.
static SFINAEResponse getDiagnosticSFINAEResponse(unsigned DiagID);
+ /// Whether the diagnostic message can be deferred.
+ ///
+ /// For single source offloading languages, a diagnostic message occurred
+ /// in a device host function may be deferred until the function is sure
+ /// to be emitted.
+ static bool isDeferrable(unsigned DiagID);
+
/// Get the string of all diagnostic flags.
///
/// \returns A list of all diagnostics flags as they would be written in a
Index: clang/include/clang/Basic/DiagnosticFrontend.h
===================================================================
--- clang/include/clang/Basic/DiagnosticFrontend.h
+++ clang/include/clang/Basic/DiagnosticFrontend.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define FRONTENDSTART
#include "clang/Basic/DiagnosticFrontendKinds.inc"
Index: clang/include/clang/Basic/DiagnosticDriver.h
===================================================================
--- clang/include/clang/Basic/DiagnosticDriver.h
+++ clang/include/clang/Basic/DiagnosticDriver.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define DRIVERSTART
#include "clang/Basic/DiagnosticDriverKinds.inc"
Index: clang/include/clang/Basic/DiagnosticCrossTU.h
===================================================================
--- clang/include/clang/Basic/DiagnosticCrossTU.h
+++ clang/include/clang/Basic/DiagnosticCrossTU.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define CROSSTUSTART
#include "clang/Basic/DiagnosticCrossTUKinds.inc"
Index: clang/include/clang/Basic/DiagnosticComment.h
===================================================================
--- clang/include/clang/Basic/DiagnosticComment.h
+++ clang/include/clang/Basic/DiagnosticComment.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define COMMENTSTART
#include "clang/Basic/DiagnosticCommentKinds.inc"
Index: clang/include/clang/Basic/DiagnosticAnalysis.h
===================================================================
--- clang/include/clang/Basic/DiagnosticAnalysis.h
+++ clang/include/clang/Basic/DiagnosticAnalysis.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define ANALYSISSTART
#include "clang/Basic/DiagnosticAnalysisKinds.inc"
Index: clang/include/clang/Basic/DiagnosticAST.h
===================================================================
--- clang/include/clang/Basic/DiagnosticAST.h
+++ clang/include/clang/Basic/DiagnosticAST.h
@@ -15,7 +15,7 @@
namespace diag {
enum {
#define DIAG(ENUM, FLAGS, DEFAULT_MAPPING, DESC, GROUP, SFINAE, NOWERROR, \
- SHOWINSYSHEADER, CATEGORY) \
+ SHOWINSYSHEADER, DEFERRABLE, CATEGORY) \
ENUM,
#define ASTSTART
#include "clang/Basic/DiagnosticASTKinds.inc"
Index: clang/include/clang/Basic/Diagnostic.td
===================================================================
--- clang/include/clang/Basic/Diagnostic.td
+++ clang/include/clang/Basic/Diagnostic.td
@@ -45,6 +45,7 @@
// diagnostics
string Component = "";
string CategoryName = "";
+ bit Deferrable = 0;
}
// Diagnostic Categories. These can be applied to groups or individual
@@ -83,6 +84,7 @@
bit AccessControl = 0;
bit WarningNoWerror = 0;
bit ShowInSystemHeader = 0;
+ bit Deferrable = 0;
Severity DefaultSeverity = defaultmapping;
DiagGroup Group;
string CategoryName = "";
@@ -106,6 +108,14 @@
bit ShowInSystemHeader = 0;
}
+class Deferrable {
+ bit Deferrable = 1;
+}
+
+class NonDeferrable {
+ bit Deferrable = 0;
+}
+
// FIXME: ExtWarn and Extension should also be SFINAEFailure by default.
class Error<string str> : Diagnostic<str, CLASS_ERROR, SEV_Error>, SFINAEFailure {
bit ShowInSystemHeader = 1;
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
