Skip to content

Commit 28ef967

Browse files
authored
[SYCL] Emit kernel args size warning for real sycl headers (#2554)
It turned out that real sycl headers are included as "system" headers and the actual warning points to function defined there, so the warning about kernel arguments size wasn't emitted.
1 parent 0c8d46e commit 28ef967

File tree

2 files changed

+22
-1
lines changed

2 files changed

+22
-1
lines changed

clang/include/clang/Basic/DiagnosticSemaKinds.td

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -11020,7 +11020,7 @@ def err_sycl_restrict : Error<
1102011020
"}0">;
1102111021
def warn_sycl_kernel_too_big_args : Warning<
1102211022
"size of kernel arguments (%0 bytes) may exceed the supported maximum "
11023-
"of %1 bytes on some devices">, InGroup<SyclStrict>;
11023+
"of %1 bytes on some devices">, InGroup<SyclStrict>, ShowInSystemHeader;
1102411024
def err_sycl_virtual_types : Error<
1102511025
"No class with a vtable can be used in a SYCL kernel or any code included in the kernel">;
1102611026
def note_sycl_recursive_function_declared_here: Note<"function implemented using recursion declared here">;
Lines changed: 21 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,21 @@
1+
// RUN: %clangxx -fsycl -fsycl-device-only -Xclang -verify -Wsycl-strict %s -fsyntax-only
2+
3+
#include <CL/sycl.hpp>
4+
5+
using namespace cl::sycl;
6+
7+
// expected-warning@../../include/sycl/CL/sycl/handler.hpp:919 {{size of kernel arguments (8068 bytes) may exceed the supported maximum of 2048 bytes on some devices}}
8+
9+
int main() {
10+
struct S {
11+
int A;
12+
int B;
13+
int Array[2015];
14+
} Args;
15+
queue myQueue;
16+
myQueue.submit([&](handler &cgh) {
17+
// expected-note@+1 {{in instantiation of function template specialization 'cl::sycl::handler::single_task}}
18+
cgh.single_task<class kernel>([=]() { (void)Args; });
19+
});
20+
return 0;
21+
}

0 commit comments

Comments
 (0)