Skip to content

Commit 208f23f

Browse files
committed
[SYCL] Improve diagnostic about invalid kernel name
Signed-off-by: Alexey Sachkov <[email protected]>
1 parent 89886da commit 208f23f

File tree

3 files changed

+14
-7
lines changed

3 files changed

+14
-7
lines changed

clang/include/clang/Basic/DiagnosticSemaKinds.td

Lines changed: 3 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -10654,8 +10654,9 @@ def err_builtin_launder_invalid_arg : Error<
1065410654
// SYCL-specific diagnostics
1065510655
def err_sycl_attribute_address_space_invalid : Error<
1065610656
"address space is outside the valid range of values">;
10657-
def err_sycl_kernel_name_class_not_top_level : Error<
10658-
"kernel needs to have a globally-visible name">;
10657+
def err_sycl_kernel_incorrectly_named : Error<
10658+
"kernel %select{name is missing"
10659+
"|needs to have a globally-visible name}0">;
1065910660
def err_sycl_restrict : Error<
1066010661
"SYCL kernel cannot "
1066110662
"%select{use a non-const global variable"

clang/lib/Sema/SemaSYCL.cpp

Lines changed: 10 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -1504,12 +1504,18 @@ void SYCLIntegrationHeader::emitFwdDecl(raw_ostream &O, const Decl *D,
15041504
if (TD && TD->isCompleteDefinition() && !UnnamedLambdaSupport) {
15051505
// defined class constituting the kernel name is not globally
15061506
// accessible - contradicts the spec
1507-
Diag.Report(KernelLocation,
1508-
diag::err_sycl_kernel_name_class_not_top_level);
1509-
if (!TD->getName().empty())
1507+
const bool KernelNameIsMissing = TD->getName().empty();
1508+
if (KernelNameIsMissing) {
1509+
Diag.Report(KernelLocation, diag::err_sycl_kernel_incorrectly_named)
1510+
<< /* kernel name is missing */ 0;
15101511
// Don't emit note if kernel name was completely omitted
1512+
} else {
1513+
Diag.Report(KernelLocation, diag::err_sycl_kernel_incorrectly_named)
1514+
<< /* kernel name is not globally-visible */ 1;
15111515
Diag.Report(D->getSourceRange().getBegin(),
1512-
diag::note_previous_decl) << TD->getName();
1516+
diag::note_previous_decl)
1517+
<< TD->getName();
1518+
}
15131519
}
15141520
}
15151521
break;

clang/test/SemaSYCL/unnamed-kernel.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -82,7 +82,7 @@ struct MyWrapper {
8282
int main() {
8383
cl::sycl::queue q;
8484
#ifndef __SYCL_UNNAMED_LAMBDA__
85-
// expected-error@+2 {{kernel needs to have a globally-visible name}}
85+
// expected-error@+2 {{kernel name is missing}}
8686
#endif
8787
q.submit([&](cl::sycl::handler &h) { h.single_task([] {}); });
8888

0 commit comments

Comments
 (0)