Skip to content

Commit 2c63209

Browse files
srividya-sundaramsrividya-sundaram
committed
Visitor to emit diagnostics
1 parent 998c97f commit 2c63209

File tree

3 files changed

+139
-52
lines changed

3 files changed

+139
-52
lines changed

‎clang/lib/Sema/SemaSYCL.cpp‎

Lines changed: 106 additions & 40 deletions
Original file line numberDiff line numberDiff line change
@@ -14,6 +14,8 @@
1414
#include "clang/AST/QualTypeNames.h"
1515
#include "clang/AST/RecordLayout.h"
1616
#include "clang/AST/RecursiveASTVisitor.h"
17+
#include "clang/AST/TemplateArgumentVisitor.h"
18+
#include "clang/AST/TypeVisitor.h"
1719
#include "clang/Analysis/CallGraph.h"
1820
#include "clang/Basic/Attributes.h"
1921
#include "clang/Basic/Builtins.h"
@@ -2473,9 +2475,111 @@ class SyclKernelIntHeaderCreator : public SyclKernelFieldHandler {
24732475

24742476
} // namespace
24752477

2478+
class SYCLTypeVisitor : public TypeVisitor<SYCLTypeVisitor>,
2479+
public ConstTemplateArgumentVisitor<SYCLTypeVisitor> {
2480+
Sema &S;
2481+
SourceLocation Loc;
2482+
using InnerTypeVisitor = TypeVisitor<SYCLTypeVisitor>;
2483+
using InnerTAVisitor = ConstTemplateArgumentVisitor<SYCLTypeVisitor>;
2484+
2485+
public:
2486+
SYCLTypeVisitor(Sema &S, SourceLocation Loc) : S(S), Loc(Loc) {}
2487+
2488+
void Visit(QualType T) {
2489+
if (T.isNull())
2490+
return;
2491+
const CXXRecordDecl *RD = T->getAsCXXRecordDecl();
2492+
if (!RD)
2493+
return;
2494+
// If KernelNameType has template args visit each template arg via
2495+
// ConstTemplateArgumentVisitor
2496+
if (const auto *TSD = dyn_cast<ClassTemplateSpecializationDecl>(RD)) {
2497+
const TemplateArgumentList &Args = TSD->getTemplateArgs();
2498+
for (unsigned I = 0; I < Args.size(); I++) {
2499+
const TemplateArgument &TemplateArg = Args[I];
2500+
Visit(TemplateArg);
2501+
}
2502+
} else {
2503+
InnerTypeVisitor::Visit(T.getTypePtr());
2504+
}
2505+
}
2506+
2507+
void Visit(const TemplateArgument &TA) { InnerTAVisitor::Visit(TA); }
2508+
2509+
void VisitEnumType(const EnumType *T) {
2510+
const EnumDecl *ED = T->getDecl();
2511+
if (!ED->isScoped() && !ED->isFixed()) {
2512+
S.Diag(Loc, diag::err_sycl_kernel_incorrectly_named) << 2;
2513+
S.Diag(ED->getSourceRange().getBegin(), diag::note_entity_declared_at)
2514+
<< ED;
2515+
return;
2516+
}
2517+
}
2518+
2519+
void VisitRecordType(const RecordType *T) {
2520+
return VisitTagDecl(T->getDecl());
2521+
}
2522+
2523+
void VisitTagDecl(const TagDecl *Tag) {
2524+
bool UnnamedLambdaEnabled =
2525+
S.getASTContext().getLangOpts().SYCLUnnamedLambda;
2526+
if (Tag && !UnnamedLambdaEnabled) {
2527+
const bool KernelNameIsMissing = Tag->getName().empty();
2528+
if (KernelNameIsMissing) {
2529+
S.Diag(Loc, diag::err_sycl_kernel_incorrectly_named)
2530+
<< /* kernel name is missing */ 0;
2531+
return;
2532+
} else {
2533+
if (Tag->isCompleteDefinition())
2534+
S.Diag(Loc, diag::err_sycl_kernel_incorrectly_named)
2535+
<< /* kernel name is not globally-visible */ 1;
2536+
else
2537+
S.Diag(Loc, diag::warn_sycl_implicit_decl);
2538+
2539+
S.Diag(Tag->getSourceRange().getBegin(), diag::note_previous_decl)
2540+
<< Tag->getName();
2541+
return;
2542+
}
2543+
}
2544+
}
2545+
2546+
void VisitTypeTemplateArgument(const TemplateArgument &TA) {
2547+
QualType T = TA.getAsType();
2548+
if (const auto *ET = T->getAs<EnumType>()) {
2549+
VisitEnumType(ET);
2550+
} else {
2551+
Visit(T);
2552+
}
2553+
return;
2554+
}
2555+
void VisitIntegralTemplateArgument(const TemplateArgument &TA) {
2556+
QualType T = TA.getIntegralType();
2557+
if (const EnumType *ET = T->getAs<EnumType>()) {
2558+
VisitEnumType(ET);
2559+
}
2560+
return;
2561+
}
2562+
2563+
void VisitTemplateTemplateArgument(const TemplateArgument &TA) {
2564+
TemplateDecl *TD = TA.getAsTemplate().getAsTemplateDecl();
2565+
TemplateParameterList *TemplateParams = TD->getTemplateParameters();
2566+
for (NamedDecl *P : *TemplateParams) {
2567+
if (NonTypeTemplateParmDecl *TemplateParam =
2568+
dyn_cast<NonTypeTemplateParmDecl>(P)) {
2569+
QualType T = TemplateParam->getType();
2570+
if (const EnumType *ET = T->getAs<EnumType>()) {
2571+
VisitEnumType(ET);
2572+
}
2573+
}
2574+
}
2575+
}
2576+
};
2577+
24762578
void Sema::CheckSYCLKernelCall(FunctionDecl *KernelFunc, SourceRange CallLoc,
24772579
ArrayRef<const Expr *> Args) {
24782580
const CXXRecordDecl *KernelObj = getKernelObjectType(KernelFunc);
2581+
QualType KernelNameType =
2582+
calculateKernelNameType(getASTContext(), KernelFunc);
24792583
if (!KernelObj) {
24802584
Diag(Args[0]->getExprLoc(), diag::err_sycl_kernel_not_function_object);
24812585
KernelFunc->setInvalidDecl();
@@ -2511,6 +2615,8 @@ void Sema::CheckSYCLKernelCall(FunctionDecl *KernelFunc, SourceRange CallLoc,
25112615
return;
25122616

25132617
KernelObjVisitor Visitor{*this};
2618+
SYCLTypeVisitor KernelTypeVisitor(*this, Args[0]->getExprLoc());
2619+
(void)KernelTypeVisitor.Visit(KernelNameType);
25142620
DiagnosingSYCLKernel = true;
25152621
Visitor.VisitRecordBases(KernelObj, FieldChecker, UnionChecker,
25162622
ArgsSizeChecker);
@@ -2856,18 +2962,6 @@ static void emitWithoutAnonNamespaces(llvm::raw_ostream &OS, StringRef Source) {
28562962
OS << Source;
28572963
}
28582964

2859-
static bool checkEnumTemplateParameter(const EnumDecl *ED,
2860-
DiagnosticsEngine &Diag,
2861-
SourceLocation KernelLocation) {
2862-
if (!ED->isScoped() && !ED->isFixed()) {
2863-
Diag.Report(KernelLocation, diag::err_sycl_kernel_incorrectly_named) << 2;
2864-
Diag.Report(ED->getSourceRange().getBegin(), diag::note_entity_declared_at)
2865-
<< ED;
2866-
return true;
2867-
}
2868-
return false;
2869-
}
2870-
28712965
// Emits a forward declaration
28722966
void SYCLIntegrationHeader::emitFwdDecl(raw_ostream &O, const Decl *D,
28732967
SourceLocation KernelLocation) {
@@ -2880,32 +2974,6 @@ void SYCLIntegrationHeader::emitFwdDecl(raw_ostream &O, const Decl *D,
28802974
auto *NS = dyn_cast_or_null<NamespaceDecl>(DC);
28812975

28822976
if (!NS) {
2883-
if (!DC->isTranslationUnit()) {
2884-
const TagDecl *TD = isa<ClassTemplateDecl>(D)
2885-
? cast<ClassTemplateDecl>(D)->getTemplatedDecl()
2886-
: dyn_cast<TagDecl>(D);
2887-
2888-
if (TD && !UnnamedLambdaSupport) {
2889-
// defined class constituting the kernel name is not globally
2890-
// accessible - contradicts the spec
2891-
const bool KernelNameIsMissing = TD->getName().empty();
2892-
if (KernelNameIsMissing) {
2893-
Diag.Report(KernelLocation, diag::err_sycl_kernel_incorrectly_named)
2894-
<< /* kernel name is missing */ 0;
2895-
// Don't emit note if kernel name was completely omitted
2896-
} else {
2897-
if (TD->isCompleteDefinition())
2898-
Diag.Report(KernelLocation,
2899-
diag::err_sycl_kernel_incorrectly_named)
2900-
<< /* kernel name is not globally-visible */ 1;
2901-
else
2902-
Diag.Report(KernelLocation, diag::warn_sycl_implicit_decl);
2903-
Diag.Report(D->getSourceRange().getBegin(),
2904-
diag::note_previous_decl)
2905-
<< TD->getName();
2906-
}
2907-
}
2908-
}
29092977
break;
29102978
}
29112979
++NamespaceCnt;
@@ -3013,7 +3081,6 @@ void SYCLIntegrationHeader::emitForwardClassDecls(
30133081
// Handle Kernel Name Type templated using enum type and value.
30143082
if (const auto *ET = T->getAs<EnumType>()) {
30153083
const EnumDecl *ED = ET->getDecl();
3016-
if (!checkEnumTemplateParameter(ED, Diag, KernelLocation))
30173084
emitFwdDecl(O, ED, KernelLocation);
30183085
} else if (Arg.getKind() == TemplateArgument::ArgKind::Type)
30193086
emitForwardClassDecls(O, T, KernelLocation, Printed);
@@ -3073,7 +3140,6 @@ void SYCLIntegrationHeader::emitForwardClassDecls(
30733140
QualType T = TemplateParam->getType();
30743141
if (const auto *ET = T->getAs<EnumType>()) {
30753142
const EnumDecl *ED = ET->getDecl();
3076-
if (!checkEnumTemplateParameter(ED, Diag, KernelLocation))
30773143
emitFwdDecl(O, ED, KernelLocation);
30783144
}
30793145
}

‎clang/test/SemaSYCL/kernelname-enum.cpp‎

Lines changed: 15 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1,13 +1,14 @@
11
// RUN: %clang_cc1 -fsycl -fsycl-is-device -fsycl-int-header=%t.h -fsyntax-only -Wno-sycl-2017-compat -verify %s
22

3+
// expected-error@Inputs/sycl.hpp:220 2{{kernel name is invalid. Unscoped enum requires fixed underlying type}}
34
#include "Inputs/sycl.hpp"
45

56
enum unscoped_enum_int : int {
67
val_1,
78
val_2
89
};
910

10-
// expected-note@+1 {{'unscoped_enum_no_type_set' declared here}}
11+
// expected-note@+1 1+ {{'unscoped_enum_no_type_set' declared here}}
1112
enum unscoped_enum_no_type_set {
1213
val_3,
1314
val_4
@@ -29,13 +30,18 @@ class dummy_functor_1 {
2930
void operator()() const {}
3031
};
3132

32-
// expected-error@+2 {{kernel name is invalid. Unscoped enum requires fixed underlying type}}
3333
template <unscoped_enum_no_type_set EnumType>
3434
class dummy_functor_2 {
3535
public:
3636
void operator()() const {}
3737
};
3838

39+
template <template <unscoped_enum_no_type_set EnumType> class C>
40+
class templated_functor {
41+
public:
42+
void operator()() const {}
43+
};
44+
3945
template <scoped_enum_int EnumType>
4046
class dummy_functor_3 {
4147
public:
@@ -54,6 +60,7 @@ int main() {
5460
dummy_functor_2<val_3> f2;
5561
dummy_functor_3<scoped_enum_int::val_2> f3;
5662
dummy_functor_4<scoped_enum_no_type_set::val_4> f4;
63+
templated_functor<dummy_functor_2> f5;
5764

5865
cl::sycl::queue q;
5966

@@ -62,9 +69,15 @@ int main() {
6269
});
6370

6471
q.submit([&](cl::sycl::handler &cgh) {
72+
// expected-note@+1{{in instantiation of function template specialization}}
6573
cgh.single_task(f2);
6674
});
6775

76+
q.submit([&](cl::sycl::handler &cgh) {
77+
// expected-note@+1{{in instantiation of function template specialization}}
78+
cgh.single_task(f5);
79+
});
80+
6881
q.submit([&](cl::sycl::handler &cgh) {
6982
cgh.single_task(f3);
7083
});

‎clang/test/SemaSYCL/unnamed-kernel.cpp‎

Lines changed: 18 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -10,7 +10,7 @@ namespace namespace1 {
1010
template <typename T>
1111
class KernelName;
1212
}
13-
13+
// expected-note@14 {{MyWrapper declared here}}
1414
struct MyWrapper {
1515
private:
1616
class InvalidKernelName0 {};
@@ -22,59 +22,66 @@ struct MyWrapper {
2222
void test() {
2323
cl::sycl::queue q;
2424
#ifndef __SYCL_UNNAMED_LAMBDA__
25-
// expected-error@+5 {{kernel needs to have a globally-visible name}}
25+
// expected-error@Inputs/sycl.hpp:220 {{kernel needs to have a globally-visible name}}
2626
// expected-note@+2 {{InvalidKernelName1 declared here}}
2727
#endif
2828
class InvalidKernelName1 {};
2929
q.submit([&](cl::sycl::handler &h) {
30+
// expected-note@+1{{in instantiation of function template specialization}}
3031
h.single_task<InvalidKernelName1>([] {});
3132
});
3233

3334
#ifndef __SYCL_UNNAMED_LAMBDA__
34-
// expected-error@+5 {{kernel needs to have a globally-visible name}}
35+
// expected-error@Inputs/sycl.hpp:220 {{kernel needs to have a globally-visible name}}
3536
// expected-note@+2 {{InvalidKernelName2 declared here}}
3637
#endif
3738
class InvalidKernelName2 {};
3839
q.submit([&](cl::sycl::handler &h) {
40+
// expected-note@+1{{in instantiation of function template specialization}}
3941
h.single_task<namespace1::KernelName<InvalidKernelName2>>([] {});
4042
});
4143

4244
#ifndef __SYCL_UNNAMED_LAMBDA__
43-
// expected-error@+4 {{kernel needs to have a globally-visible name}}
45+
// expected-error@Inputs/sycl.hpp:220 {{kernel needs to have a globally-visible name}}
4446
// expected-note@16 {{InvalidKernelName0 declared here}}
4547
#endif
4648
q.submit([&](cl::sycl::handler &h) {
49+
// expected-note@+1{{in instantiation of function template specialization}}
4750
h.single_task<InvalidKernelName0>([] {});
4851
});
4952

5053
#ifndef __SYCL_UNNAMED_LAMBDA__
51-
// expected-error@+4 {{kernel needs to have a globally-visible name}}
54+
// expected-error@Inputs/sycl.hpp:220 {{kernel needs to have a globally-visible name}}
5255
// expected-note@17 {{InvalidKernelName3 declared here}}
5356
#endif
5457
q.submit([&](cl::sycl::handler &h) {
58+
// expected-note@+1{{in instantiation of function template specialization}}
5559
h.single_task<namespace1::KernelName<InvalidKernelName3>>([] {});
5660
});
5761

58-
using ValidAlias = MyWrapper;
62+
/* using ValidAlias = MyWrapper;
5963
q.submit([&](cl::sycl::handler &h) {
64+
6065
h.single_task<ValidAlias>([] {});
61-
});
66+
}); */
6267

6368
using InvalidAlias = InvalidKernelName4;
6469
#ifndef __SYCL_UNNAMED_LAMBDA__
65-
// expected-error@+4 {{kernel needs to have a globally-visible name}}
70+
// expected-error@Inputs/sycl.hpp:220 {{kernel needs to have a globally-visible name}}
6671
// expected-note@18 {{InvalidKernelName4 declared here}}
6772
#endif
6873
q.submit([&](cl::sycl::handler &h) {
74+
// expected-note@+1{{in instantiation of function template specialization}}
6975
h.single_task<InvalidAlias>([] {});
7076
});
7177

7278
using InvalidAlias1 = InvalidKernelName5;
7379
#ifndef __SYCL_UNNAMED_LAMBDA__
74-
// expected-error@+4 {{kernel needs to have a globally-visible name}}
80+
// expected-error@Inputs/sycl.hpp:220 {{kernel needs to have a globally-visible name}}
7581
// expected-note@19 {{InvalidKernelName5 declared here}}
7682
#endif
7783
q.submit([&](cl::sycl::handler &h) {
84+
// expected-note@+1{{in instantiation of function template specialization}}
7885
h.single_task<namespace1::KernelName<InvalidAlias1>>([] {});
7986
});
8087
}
@@ -83,8 +90,9 @@ struct MyWrapper {
8390
int main() {
8491
cl::sycl::queue q;
8592
#ifndef __SYCL_UNNAMED_LAMBDA__
86-
// expected-error@+2 {{kernel name is missing}}
93+
// expected-error@Inputs/sycl.hpp:220 {{kernel name is missing}}
8794
#endif
95+
// expected-note@+1{{in instantiation of function template specialization}}
8896
q.submit([&](cl::sycl::handler &h) { h.single_task([] {}); });
8997

9098
return 0;

0 commit comments

Comments
 (0)