Skip to content

Commit 55f51e3

Browse files
committed
[SYCL] Emit kernel args size warning for real sycl headers
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 7b4672d commit 55f51e3

File tree

2 files changed

+23
-1
lines changed

2 files changed

+23
-1
lines changed

clang/include/clang/Basic/DiagnosticSemaKinds.td

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

0 commit comments

Comments
 (0)