From 815e6c4cc7b02ee3953216fa4362c8b60f2e6365 Mon Sep 17 00:00:00 2001 From: Erich Keane Date: Fri, 7 May 2021 10:34:21 -0700 Subject: [PATCH 1/4] [SYCL] Don't diagnose checks in a discarded statement. For the checks we do in SemaSYCL via the callgraph, we don't want to diagnose when it is in a discarded statement. This adds the infrastructure for another bug, where we visit constexpr during this too. --- clang/include/clang/AST/Stmt.h | 1 + clang/include/clang/Analysis/CallGraph.h | 19 ++++++++++++++++++- clang/lib/AST/Stmt.cpp | 6 ++++++ clang/lib/Analysis/CallGraph.cpp | 14 ++++++++++++++ clang/lib/Sema/SemaSYCL.cpp | 1 + 5 files changed, 40 insertions(+), 1 deletion(-) diff --git a/clang/include/clang/AST/Stmt.h b/clang/include/clang/AST/Stmt.h index 97cd903f3d6f4..258b17e83b881 100644 --- a/clang/include/clang/AST/Stmt.h +++ b/clang/include/clang/AST/Stmt.h @@ -2080,6 +2080,7 @@ class IfStmt final /// If this is an 'if constexpr', determine which substatement will be taken. /// Otherwise, or if the condition is value-dependent, returns None. Optional getNondiscardedCase(const ASTContext &Ctx) const; + Optional getNondiscardedCase(const ASTContext &Ctx); bool isObjCAvailabilityCheck() const; diff --git a/clang/include/clang/Analysis/CallGraph.h b/clang/include/clang/Analysis/CallGraph.h index 999ac5da8acb6..4b4344814ba65 100644 --- a/clang/include/clang/Analysis/CallGraph.h +++ b/clang/include/clang/Analysis/CallGraph.h @@ -29,6 +29,7 @@ namespace clang { +class ASTContext; class CallGraphNode; class Decl; class DeclContext; @@ -51,6 +52,12 @@ class CallGraph : public RecursiveASTVisitor { /// This is a virtual root node that has edges to all the functions. CallGraphNode *Root; + /// A setting to determine whether this should include calls that are done in + /// a constant expression's context. This DOES require the ASTContext object + /// for constexpr-if, so setting it requires a valid ASTContext. + bool shouldSkipConstexpr = false; + ASTContext *Ctx; + public: CallGraph(); ~CallGraph(); @@ -66,7 +73,7 @@ class CallGraph : public RecursiveASTVisitor { /// Determine if a declaration should be included in the graph. static bool includeInGraph(const Decl *D); - /// Determine if a declaration should be included in the graph for the + /// Determine if a declaration should be included in the graph for the /// purposes of being a callee. This is similar to includeInGraph except /// it permits declarations, not just definitions. static bool includeCalleeInGraph(const Decl *D); @@ -138,6 +145,16 @@ class CallGraph : public RecursiveASTVisitor { bool shouldWalkTypesOfTypeLocs() const { return false; } bool shouldVisitTemplateInstantiations() const { return true; } bool shouldVisitImplicitCode() const { return true; } + bool shouldVisitConstantExpressions() const { return false; } + bool shouldSkipConstantExpressions() const { return shouldSkipConstexpr; } + void setSkipConstantExpressions(ASTContext &Context) { + Ctx = &Context; + shouldSkipConstexpr = true; + } + ASTContext &getASTContext() { + assert(Ctx); + return *Ctx; + } private: /// Add the given declaration to the call graph. diff --git a/clang/lib/AST/Stmt.cpp b/clang/lib/AST/Stmt.cpp index 2ceee614cf98f..8073d7bec19d8 100644 --- a/clang/lib/AST/Stmt.cpp +++ b/clang/lib/AST/Stmt.cpp @@ -995,6 +995,12 @@ Optional IfStmt::getNondiscardedCase(const ASTContext &Ctx) const { return !getCond()->EvaluateKnownConstInt(Ctx) ? getElse() : getThen(); } +Optional IfStmt::getNondiscardedCase(const ASTContext &Ctx) { + if (!isConstexpr() || getCond()->isValueDependent()) + return None; + return !getCond()->EvaluateKnownConstInt(Ctx) ? getElse() : getThen(); +} + ForStmt::ForStmt(const ASTContext &C, Stmt *Init, Expr *Cond, VarDecl *condVar, Expr *Inc, Stmt *Body, SourceLocation FL, SourceLocation LP, SourceLocation RP) diff --git a/clang/lib/Analysis/CallGraph.cpp b/clang/lib/Analysis/CallGraph.cpp index 2299ba32db501..6f120759e0df8 100644 --- a/clang/lib/Analysis/CallGraph.cpp +++ b/clang/lib/Analysis/CallGraph.cpp @@ -11,6 +11,7 @@ //===----------------------------------------------------------------------===// #include "clang/Analysis/CallGraph.h" +#include "clang/AST/ASTContext.h" #include "clang/AST/Decl.h" #include "clang/AST/DeclBase.h" #include "clang/AST/DeclObjC.h" @@ -136,6 +137,19 @@ class CGBuilder : public StmtVisitor { } } + void VisitIfStmt(IfStmt *If) { + if (G->shouldSkipConstantExpressions()) { + if (llvm::Optional ActiveStmt = + If->getNondiscardedCase(G->getASTContext())) { + if (*ActiveStmt) + this->Visit(*ActiveStmt); + return; + } + } + + VisitChildren(If); + } + void VisitChildren(Stmt *S) { for (Stmt *SubStmt : S->children()) if (SubStmt) diff --git a/clang/lib/Sema/SemaSYCL.cpp b/clang/lib/Sema/SemaSYCL.cpp index 7cfb5bdbc5b7a..32c5584ecf4c9 100644 --- a/clang/lib/Sema/SemaSYCL.cpp +++ b/clang/lib/Sema/SemaSYCL.cpp @@ -548,6 +548,7 @@ class DeviceFunctionTracker { public: DeviceFunctionTracker(Sema &S) : SemaRef(S) { + CG.setSkipConstantExpressions(S.Context); CG.addToCallGraph(S.getASTContext().getTranslationUnitDecl()); CollectSyclExternalFuncs(); } From 25a9fa3181cc6b38e9bd1e6c39b53c6f3fae541d Mon Sep 17 00:00:00 2001 From: Erich Keane Date: Wed, 12 May 2021 10:26:29 -0700 Subject: [PATCH 2/4] Stop visiting the initializers of constexpr vardecls. --- clang/lib/Analysis/CallGraph.cpp | 20 +++++++++++++++++++- 1 file changed, 19 insertions(+), 1 deletion(-) diff --git a/clang/lib/Analysis/CallGraph.cpp b/clang/lib/Analysis/CallGraph.cpp index 6f120759e0df8..78197ea5e65c2 100644 --- a/clang/lib/Analysis/CallGraph.cpp +++ b/clang/lib/Analysis/CallGraph.cpp @@ -147,7 +147,25 @@ class CGBuilder : public StmtVisitor { } } - VisitChildren(If); + StmtVisitor::VisitIfStmt(If); + } + + void VisitDeclStmt(DeclStmt *DS) { + if (G->shouldSkipConstantExpressions()) { + auto IsConstexprVarDecl = [](Decl *D) { + if (const auto *VD = dyn_cast(D)) + return VD->isConstexpr(); + return false; + }; + if (llvm::any_of(DS->decls(), IsConstexprVarDecl)) { + assert(llvm::all_of(DS->decls(), IsConstexprVarDecl) && + "Situation where a decl-group would be a mix of decl types, or " + "constexpr and not?"); + return; + } + } + + StmtVisitor::VisitDeclStmt(DS); } void VisitChildren(Stmt *S) { From 6fd33a9128bb70d5c7f29a1a01fbb74abab4e12b Mon Sep 17 00:00:00 2001 From: Elizabeth Andrews Date: Fri, 18 Jun 2021 11:31:06 -0700 Subject: [PATCH 3/4] [SYCL] Fix bugs with recursion in SYCL kernel This patch removes ConstexprDepthRAII which incorrectly maintained constexpr context. Instead we don't traverse statements which are forced constant expressions. This patch fixes the following bugs - 1. constexpr int j = test_constexpr_context(recfn(1)); 2. int k; if constexpr (false) k = recfn(1); Here, recfn() is a recursive function. The usage 1. and 2. should not diagnose in SYCL kernel since the function is called in constexpr context. Signed-off-by: Elizabeth Andrews --- clang/include/clang/Analysis/CallGraph.h | 1 - clang/lib/Sema/SemaSYCL.cpp | 52 ++++++------------- .../SemaSYCL/allow-constexpr-recursion.cpp | 42 ++++++++++++--- 3 files changed, 50 insertions(+), 45 deletions(-) diff --git a/clang/include/clang/Analysis/CallGraph.h b/clang/include/clang/Analysis/CallGraph.h index 4b4344814ba65..0c98b1ca354b4 100644 --- a/clang/include/clang/Analysis/CallGraph.h +++ b/clang/include/clang/Analysis/CallGraph.h @@ -145,7 +145,6 @@ class CallGraph : public RecursiveASTVisitor { bool shouldWalkTypesOfTypeLocs() const { return false; } bool shouldVisitTemplateInstantiations() const { return true; } bool shouldVisitImplicitCode() const { return true; } - bool shouldVisitConstantExpressions() const { return false; } bool shouldSkipConstantExpressions() const { return shouldSkipConstexpr; } void setSkipConstantExpressions(ASTContext &Context) { Ctx = &Context; diff --git a/clang/lib/Sema/SemaSYCL.cpp b/clang/lib/Sema/SemaSYCL.cpp index 449047239bea0..a3af2963a7300 100644 --- a/clang/lib/Sema/SemaSYCL.cpp +++ b/clang/lib/Sema/SemaSYCL.cpp @@ -395,27 +395,9 @@ static void collectSYCLAttributes(Sema &S, FunctionDecl *FD, } class DiagDeviceFunction : public RecursiveASTVisitor { - // Used to keep track of the constexpr depth, so we know whether to skip - // diagnostics. - unsigned ConstexprDepth = 0; Sema &SemaRef; const llvm::SmallPtrSetImpl &RecursiveFuncs; - struct ConstexprDepthRAII { - DiagDeviceFunction &DDF; - bool Increment; - - ConstexprDepthRAII(DiagDeviceFunction &DDF, bool Increment = true) - : DDF(DDF), Increment(Increment) { - if (Increment) - ++DDF.ConstexprDepth; - } - ~ConstexprDepthRAII() { - if (Increment) - --DDF.ConstexprDepth; - } - }; - public: DiagDeviceFunction( Sema &S, @@ -433,7 +415,7 @@ class DiagDeviceFunction : public RecursiveASTVisitor { // instantiation as template functions. It means that // all functions used by kernel have already been parsed and have // definitions. - if (RecursiveFuncs.count(Callee) && !ConstexprDepth) { + if (RecursiveFuncs.count(Callee)) { SemaRef.Diag(e->getExprLoc(), diag::err_sycl_restrict) << Sema::KernelCallRecursiveFunction; SemaRef.Diag(Callee->getSourceRange().getBegin(), @@ -486,45 +468,41 @@ class DiagDeviceFunction : public RecursiveASTVisitor { // Skip checking rules on variables initialized during constant evaluation. bool TraverseVarDecl(VarDecl *VD) { - ConstexprDepthRAII R(*this, VD->isConstexpr()); + if (VD->isConstexpr()) + return true; return RecursiveASTVisitor::TraverseVarDecl(VD); } // Skip checking rules on template arguments, since these are constant // expressions. bool TraverseTemplateArgumentLoc(const TemplateArgumentLoc &ArgLoc) { - ConstexprDepthRAII R(*this); - return RecursiveASTVisitor::TraverseTemplateArgumentLoc(ArgLoc); + return true; } // Skip checking the static assert, both components are required to be // constant expressions. - bool TraverseStaticAssertDecl(StaticAssertDecl *D) { - ConstexprDepthRAII R(*this); - return RecursiveASTVisitor::TraverseStaticAssertDecl(D); - } + bool TraverseStaticAssertDecl(StaticAssertDecl *D) { return true; } // Make sure we skip the condition of the case, since that is a constant // expression. bool TraverseCaseStmt(CaseStmt *S) { - { - ConstexprDepthRAII R(*this); - if (!TraverseStmt(S->getLHS())) - return false; - if (!TraverseStmt(S->getRHS())) - return false; - } return TraverseStmt(S->getSubStmt()); } // Skip checking the size expr, since a constant array type loc's size expr is // a constant expression. bool TraverseConstantArrayTypeLoc(const ConstantArrayTypeLoc &ArrLoc) { - if (!TraverseTypeLoc(ArrLoc.getElementLoc())) - return false; + return true; + } - ConstexprDepthRAII R(*this); - return TraverseStmt(ArrLoc.getSizeExpr()); + bool TraverseIfStmt(IfStmt *S) { + if (llvm::Optional ActiveStmt = + S->getNondiscardedCase(SemaRef.Context)) { + if (*ActiveStmt) + return TraverseStmt(*ActiveStmt); + return true; + } + return RecursiveASTVisitor::TraverseIfStmt(S); } }; diff --git a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp index 4dde146900849..0e888eb15d888 100644 --- a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp +++ b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp @@ -8,7 +8,7 @@ sycl::queue q; constexpr int constexpr_recurse1(int n); -// expected-note@+1 3{{function implemented using recursion declared here}} +// expected-note@+1 5{{function implemented using recursion declared here}} constexpr int constexpr_recurse(int n) { if (n) return constexpr_recurse1(n - 1); @@ -20,6 +20,10 @@ constexpr int constexpr_recurse1(int n) { return constexpr_recurse(n) + 1; } +constexpr int test_constexpr_context(int n) { + return n; +} + template void bar() {} @@ -55,15 +59,13 @@ void ConstexprIf2() { // they should not diagnose. void constexpr_recurse_test() { constexpr int i = constexpr_recurse(1); + constexpr int j = test_constexpr_context(constexpr_recurse(1)); bar(); bar2<1, 2, constexpr_recurse(2)>(); static_assert(constexpr_recurse(2) == 105, ""); - int j; switch (105) { case constexpr_recurse(2): - // expected-error@+1{{SYCL kernel cannot call a recursive function}} - j = constexpr_recurse(5); break; } @@ -78,14 +80,40 @@ void constexpr_recurse_test() { ConditionallyExplicitCtor c(1); - ConstexprIf1<0>(); // Should not cause a diagnostic. - // expected-error@+1{{SYCL kernel cannot call a recursive function}} - ConstexprIf2<1>(); + ConstexprIf1<0>(); + + int k; + if constexpr (false) + k = constexpr_recurse(1); + else + constexpr int l = test_constexpr_context(constexpr_recurse(1)); } void constexpr_recurse_test_err() { // expected-error@+1{{SYCL kernel cannot call a recursive function}} int i = constexpr_recurse(1); + + // expected-error@+1{{SYCL kernel cannot call a recursive function}} + ConstexprIf2<1>(); + + int j, k; + if constexpr (true) + // expected-error@+1{{SYCL kernel cannot call a recursive function}} + j = constexpr_recurse(1); + + if constexpr (false) + j = constexpr_recurse(1); // Should not diagnose in discarded branch + else + // expected-error@+1{{SYCL kernel cannot call a recursive function}} + k = constexpr_recurse(1); + + switch (105) { + case constexpr_recurse(2): + constexpr int l = test_constexpr_context(constexpr_recurse(1)); + // expected-error@+1{{SYCL kernel cannot call a recursive function}} + j = constexpr_recurse(5); + break; + } } int main() { From be2739c7dc017b6469d3d132ad602355284ba1dc Mon Sep 17 00:00:00 2001 From: Elizabeth Andrews Date: Mon, 21 Jun 2021 06:59:10 -0700 Subject: [PATCH 4/4] Implement review comment - fix variable name Signed-off-by: Elizabeth Andrews --- clang/include/clang/Analysis/CallGraph.h | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/clang/include/clang/Analysis/CallGraph.h b/clang/include/clang/Analysis/CallGraph.h index 0c98b1ca354b4..a63ded97dd030 100644 --- a/clang/include/clang/Analysis/CallGraph.h +++ b/clang/include/clang/Analysis/CallGraph.h @@ -55,7 +55,7 @@ class CallGraph : public RecursiveASTVisitor { /// A setting to determine whether this should include calls that are done in /// a constant expression's context. This DOES require the ASTContext object /// for constexpr-if, so setting it requires a valid ASTContext. - bool shouldSkipConstexpr = false; + bool ShouldSkipConstexpr = false; ASTContext *Ctx; public: @@ -145,10 +145,10 @@ class CallGraph : public RecursiveASTVisitor { bool shouldWalkTypesOfTypeLocs() const { return false; } bool shouldVisitTemplateInstantiations() const { return true; } bool shouldVisitImplicitCode() const { return true; } - bool shouldSkipConstantExpressions() const { return shouldSkipConstexpr; } + bool shouldSkipConstantExpressions() const { return ShouldSkipConstexpr; } void setSkipConstantExpressions(ASTContext &Context) { Ctx = &Context; - shouldSkipConstexpr = true; + ShouldSkipConstexpr = true; } ASTContext &getASTContext() { assert(Ctx);