Skip to content

[SYCL] Reject zero length arrays #1153

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 19 commits into from
Mar 3, 2020
Merged

[SYCL] Reject zero length arrays #1153

Show file tree
Hide file tree
Changes from 9 commits
Commits
Show all changes
19 commits
Select commit Hold shift + click to select a range
eaa10ea
disallow zero length arrays in SYCL kernels
cperkinsintel Feb 15, 2020
e9dd04a
changing diagnostic code per feedback
cperkinsintel Feb 20, 2020
765d43b
remove braces, per PR request
cperkinsintel Feb 20, 2020
fb3d2ac
adding test of deferred diagnostics, per PR request
cperkinsintel Feb 20, 2020
9618333
improved deferred diagnostic testing
cperkinsintel Feb 20, 2020
cc9aa9f
Update clang/test/SemaSYCL/deferred-diagnostics-emit.cpp
cperkinsintel Feb 21, 2020
8d92c4d
switched to common fake SYCL scaffolding
cperkinsintel Feb 21, 2020
17a2f44
Update clang/lib/Sema/SemaType.cpp
cperkinsintel Feb 21, 2020
b8a67d3
changing comment formatting per PR feedback
cperkinsintel Feb 21, 2020
e28c09e
Update clang/test/SemaSYCL/deferred-diagnostics-emit.cpp
cperkinsintel Feb 21, 2020
8634c6e
Update clang/test/SemaSYCL/deferred-diagnostics-emit.cpp
cperkinsintel Feb 25, 2020
0a208a1
updating test. Having problem with 'note' expectation.
cperkinsintel Feb 25, 2020
b515e70
Merge branch 'cperkins-sycl-reject-zero-length-arrays' of https://git…
cperkinsintel Feb 25, 2020
a1c2eb3
llvm-lit test passing
cperkinsintel Feb 26, 2020
11b2e94
clang-format
cperkinsintel Feb 26, 2020
7d7a2ec
quick comm
cperkinsintel Feb 26, 2020
77a5c88
cf trouble
cperkinsintel Feb 27, 2020
f63f410
comment
cperkinsintel Feb 27, 2020
dd87de2
lightening test
cperkinsintel Feb 28, 2020
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
6 changes: 6 additions & 0 deletions clang/lib/Sema/SemaType.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -2340,6 +2340,12 @@ QualType Sema::BuildArrayType(QualType T, ArrayType::ArraySizeModifier ASM,
<< ArraySize->getSourceRange();
ASM = ArrayType::Normal;
}

// Zero length arrays are disallowed in SYCL device code.
if (getLangOpts().SYCLIsDevice)
SYCLDiagIfDeviceCode(ArraySize->getBeginLoc(),
diag::err_typecheck_zero_array_size)
<< ArraySize->getSourceRange();
} else if (!T->isDependentType() && !T->isVariablyModifiedType() &&
!T->isIncompleteType() && !T->isUndeducedType()) {
// Is the array too large?
Expand Down
85 changes: 85 additions & 0 deletions clang/test/SemaSYCL/deferred-diagnostics-emit.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,85 @@
// RUN: %clang_cc1 -I %S/Inputs -fcxx-exceptions -triple spir64 -fsycl-is-device -Wno-return-type -verify -fsyntax-only %s

//
// ensuring that the SYCL diagnostics that are typically deferred, correctly emit
//

#include <sycl.hpp>

struct S {
virtual void foo() {}
};

int calledFromKernel(int a){
// expected-error@+1 {{zero-length arrays are not permitted in C++}}
int MalArray[0];

// expected-error@+1 {{__float128 is not supported on this target}}
__float128 malFloat = 40;

// not sure if 'no virtual function' is a _deferred_ diagnostic, testing anyway
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

What exactly do you mean here?
BTW, we diagnose only usages of virtual functions, as SYCL spec requires. Using of classes with virtual functions is not prohibited.

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I am not sure if this particular violation ( a virtual function in SYCL context ) results in a deferred diagnostic, as opposed to an immediate diagnostic. @erichkeane added it to one of the files when he was working on reducing the problem. So I kept it in the test.

Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

AFAIK we diagnose virtual functions usage through SemaSYCL callgraph traversing. See

llvm/clang/lib/Sema/SemaSYCL.cpp

Line 233 in f87291d

<< Sema::KernelCallVirtualFunction;
. But, I suppose, we will move this to Sema and deferred diagnostics.

But we already diagnose variadic function calls through deferred diagnostic, I remember patch from Erich doing this #998 .
So If here you are trying to check only deferred diagnostics, I suggest switch from virtual functions to variadic functions.

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Ok, I switched to variadic.

S mal;
// expected-error@+1 {{SYCL kernel cannot call a virtual function}}
mal.foo();

return a + 20;
}


// template used to specialize a function that contains a lambda that should
// result in a deferred diagnostic being emitted.
// HOWEVER, this is not working presently.
// TODO: re-test after new deferred diagnostic system is merged.
// restore the "FIX!!" tests below

template <typename T>
void setup_sycl_operation(const T VA[]) {

cl::sycl::range<1> numOfItems;
cl::sycl::queue deviceQueue;

deviceQueue.submit([&](cl::sycl::handler &cgh) {
cgh.single_task<class AName>([=]() {
// FIX!! expected-error@+1 {{zero-length arrays are not permitted in C++}}
int OverlookedBadArray[0];

// FIX!! expected-error@+1 {{__float128 is not supported on this target}}
__float128 overlookedBadFloat = 40;

});
});
}

int main(int argc, char **argv) {

// --- direct lambda testing ---
cl::sycl::range<1> numOfItems;
cl::sycl::queue deviceQueue;

deviceQueue.submit([&](cl::sycl::handler &cgh) {
cgh.single_task<class AName>([=]() {
Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

@Fznamznon
Can you help with with this test? I have one last "note" I need to suppress.

When running llvm-lit -v deferred-diagnostics-emit.cpp I get this:

error: 'note' diagnostics seen but not expected: 
  File /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/Inputs/sycl.hpp Line 182: called by 'kernel_single_task<AName, (lambda at /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/deferred-diagnostics-emit.cpp:61:34)>'
  File /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/Inputs/sycl.hpp Line 182: called by 'kernel_single_task<AName, (lambda at /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/deferred-diagnostics-emit.cpp:61:34)>'

So it looks like I should add this expectation:

// expected-note@+1 2{{called by 'kernel_single_task<AName, (lambda}}

But no matter where I put that expectation it never works. First it tells me it expected the note but didn't see it, then it says it saw the note without expecting it. Awesome.

error: 'note' diagnostics expected but not seen: 
  File /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/deferred-diagnostics-emit.cpp Line 61 (directive at /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/deferred-diagnostics-emit.cpp:60): called by 'kernel_single_task<AName, (lambda
  File /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/deferred-diagnostics-emit.cpp Line 61 (directive at /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/deferred-diagnostics-emit.cpp:60): called by 'kernel_single_task<AName, (lambda
error: 'note' diagnostics seen but not expected: 
  File /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/Inputs/sycl.hpp Line 182: called by 'kernel_single_task<AName, (lambda at /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/deferred-diagnostics-emit.cpp:61:34)>'
  File /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/Inputs/sycl.hpp Line 182: called by 'kernel_single_task<AName, (lambda at /iusers/cperkins/sycl_workspace/llvm/clang/test/SemaSYCL/deferred-diagnostics-emit.cpp:61:34)>'

Is the expectation correct? I haven't been able to find docs for them, so I'm just cribbing from other places.

I tried putting the expectation at line 58, line 60, in the argument list of cgh.single_task so at to precede the lambda, line 62, line 76, etc. But nothing works.

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Is there a compiler flag to suppress the notes?

Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

First, I don't think that we should suppress notes in tests.
Second, I see what happens. You are trying to add note expectation to wrong place.
See logs, the emitted note points to function named kernel_single_task and it's defined in file test/SemaSYCL/Inputs/sycl.hpp on line 182.
By the construction like // expected-note@+1 2{{called by 'kernel_single_task<AName, (lambda}} in test/SemaSYCL/deferred-diagnostics-emit.cpp file on line 60 means that you expect 2 same notes with this text in this file on line 61.
The trick here that you use single_task function from our fake sycl runtime, and kernel_single_task call happens inside it. I'm not sure that I know how to expect notes from other file (maybe @erichkeane knows). As a fix you can declare your own kernel_single_task function in your test file and note expectation will point to correct file.

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Ok - I re-introduced the bespoke SYCL namespace code into the test and put the expectation into it. Now the llvm-lit passes and the test correctly stresses the namespace aspect we require.

// expected-error@+1 {{zero-length arrays are not permitted in C++}}
int BadArray[0];

// expected-error@+1 {{__float128 is not supported on this target}}
__float128 badFloat = 40; // this SHOULD trigger a diagnostic

// not sure if 'no virtual function' is a _deferred_ diagnostic, but testing anyway.
S s;
// expected-error@+1 {{SYCL kernel cannot call a virtual function}}
s.foo();

calledFromKernel(10);
});
});


// --- lambda in specialized function testing ---

//array A is only used to feed the template
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Why this array A is needed at all?
The problem with diagnostics appears only if single_task function is called from a template function?

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Correct. The diagnostics do not emit if the lambda that contains the violations appears in a templated function. So I'm just using the array to specialize the template.

const int array_size = 4;
int A[array_size] = {1, 2, 3, 4};
setup_sycl_operation(A);

return 0;
}
3 changes: 3 additions & 0 deletions clang/test/SemaSYCL/sycl-restrict.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -143,6 +143,9 @@ void usage(myFuncDef functionPtr) {

// expected-error@+1 {{__float128 is not supported on this target}}
__float128 A;

// expected-error@+1 {{zero-length arrays are not permitted in C++}}
int BadArray[0];
}

namespace ns {
Expand Down