diff --git a/clang/test/SemaSYCL/Inputs/sycl.hpp b/clang/test/SemaSYCL/Inputs/sycl.hpp index f2bfe5e357041..6162171e84001 100644 --- a/clang/test/SemaSYCL/Inputs/sycl.hpp +++ b/clang/test/SemaSYCL/Inputs/sycl.hpp @@ -3,7 +3,7 @@ // Shared code for SYCL tests -namespace cl { +inline namespace cl { namespace sycl { namespace access { diff --git a/clang/test/SemaSYCL/accessor_inheritance.cpp b/clang/test/SemaSYCL/accessor_inheritance.cpp index 57b3179fb997f..4d445de173799 100644 --- a/clang/test/SemaSYCL/accessor_inheritance.cpp +++ b/clang/test/SemaSYCL/accessor_inheritance.cpp @@ -1,49 +1,59 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s -#include "Inputs/sycl.hpp" +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s -struct Base { +// This test checks inheritance support for struct types with accessors +// passed as kernel arguments, which are decomposed to individual fields. + +#include "sycl.hpp" + +sycl::queue myQueue; + +struct AccessorBase { int A, B; - cl::sycl::accessor AccField; + sycl::accessor AccField; }; -struct Captured : Base, - cl::sycl::accessor { +struct AccessorDerived : AccessorBase, + sycl::accessor { int C; }; int main() { - Captured Obj; - cl::sycl::kernel_single_task( - [=]() { - Obj.use(); - }); + AccessorDerived DerivedObject; + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + DerivedObject.use(); + }); + }); + + return 0; } // Check kernel parameters -// CHECK: FunctionDecl {{.*}}kernel{{.*}} 'void (int, int, __global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, __global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, int)' +// CHECK: FunctionDecl {{.*}}kernel{{.*}} 'void (int, int, __global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, __global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, int)' // CHECK: ParmVarDecl{{.*}} used _arg_A 'int' // CHECK: ParmVarDecl{{.*}} used _arg_B 'int' // CHECK: ParmVarDecl{{.*}} used _arg_AccField '__global char *' -// CHECK: ParmVarDecl{{.*}} used _arg_AccField 'cl::sycl::range<1>' -// CHECK: ParmVarDecl{{.*}} used _arg_AccField 'cl::sycl::range<1>' -// CHECK: ParmVarDecl{{.*}} used _arg_AccField 'cl::sycl::id<1>' +// CHECK: ParmVarDecl{{.*}} used _arg_AccField 'sycl::range<1>' +// CHECK: ParmVarDecl{{.*}} used _arg_AccField 'sycl::range<1>' +// CHECK: ParmVarDecl{{.*}} used _arg_AccField 'sycl::id<1>' // CHECK: ParmVarDecl{{.*}} used _arg__base '__global char *' -// CHECK: ParmVarDecl{{.*}} used _arg__base 'cl::sycl::range<1>' -// CHECK: ParmVarDecl{{.*}} used _arg__base 'cl::sycl::range<1>' -// CHECK: ParmVarDecl{{.*}} used _arg__base 'cl::sycl::id<1>' +// CHECK: ParmVarDecl{{.*}} used _arg__base 'sycl::range<1>' +// CHECK: ParmVarDecl{{.*}} used _arg__base 'sycl::range<1>' +// CHECK: ParmVarDecl{{.*}} used _arg__base 'sycl::id<1>' // CHECK: ParmVarDecl{{.*}} used _arg_C 'int' // Check lambda initialization // CHECK: VarDecl {{.*}} used '(lambda at {{.*}}accessor_inheritance.cpp // CHECK-NEXT: InitListExpr {{.*}} -// CHECK-NEXT: InitListExpr {{.*}} 'Captured' -// CHECK-NEXT: InitListExpr {{.*}} 'Base' +// CHECK-NEXT: InitListExpr {{.*}} 'AccessorDerived' +// CHECK-NEXT: InitListExpr {{.*}} 'AccessorBase' // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' // CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_A' 'int' // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' // CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_B' 'int' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::accessor':'cl::sycl::accessor' 'void () noexcept' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::accessor':'cl::sycl::accessor' 'void () noexcept' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::accessor':'sycl::accessor' 'void () noexcept' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::accessor':'sycl::accessor' 'void () noexcept' // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' // CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_C' 'int' @@ -51,15 +61,15 @@ int main() { // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} .__init // CHECK-NEXT: MemberExpr {{.*}} .AccField -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'Base' lvalue -// CHECK-NEXT: MemberExpr {{.*}} 'Captured' lvalue . +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'AccessorBase' lvalue +// CHECK-NEXT: MemberExpr {{.*}} 'AccessorDerived' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}}'(lambda at {{.*}}accessor_inheritance.cpp // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global char *' // CHECK-NEXT: DeclRefExpr {{.*}} '__global char *' lvalue ParmVar {{.*}} '_arg_AccField' '__global char *' // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr{{.*}} lvalue .__init -// CHECK-NEXT: MemberExpr{{.*}}'Captured' lvalue . +// CHECK-NEXT: MemberExpr{{.*}}'AccessorDerived' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at {{.*}}accessor_inheritance.cpp // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global char *' // CHECK-NEXT: DeclRefExpr {{.*}} '__global char *' lvalue ParmVar {{.*}} '_arg__base' '__global char *' diff --git a/clang/test/SemaSYCL/accessors-targets-image.cpp b/clang/test/SemaSYCL/accessors-targets-image.cpp index 5afa2010451cb..c6d5a19168a84 100644 --- a/clang/test/SemaSYCL/accessors-targets-image.cpp +++ b/clang/test/SemaSYCL/accessors-targets-image.cpp @@ -1,66 +1,75 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s -// This test checks that compiler generates correct kernel wrapper arguments for +// This test checks if the compiler generates correct kernel wrapper arguments for // image accessors targets. -#include "Inputs/sycl.hpp" +#include "sycl.hpp" -using namespace cl::sycl; - -template -__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) { - kernelFunc(); -} +sycl::queue q; int main() { - accessor + sycl::accessor image_acc1d_read; - kernel( - [=]() { - image_acc1d_read.use(); - }); - accessor + q.submit([&](sycl::handler &h) { + h.single_task( + [=] { + image_acc1d_read.use(); + }); + }); + + sycl::accessor image_acc2d_read; - kernel( - [=]() { - image_acc2d_read.use(); - }); + q.submit([&](sycl::handler &h) { + h.single_task( + [=] { + image_acc2d_read.use(); + }); + }); - accessor + sycl::accessor image_acc3d_read; - kernel( - [=]() { - image_acc3d_read.use(); - }); - accessor + q.submit([&](sycl::handler &h) { + h.single_task( + [=] { + image_acc3d_read.use(); + }); + }); + + sycl::accessor image_acc1d_write; - kernel( - [=]() { - image_acc1d_write.use(); - }); + q.submit([&](sycl::handler &h) { + h.single_task( + [=] { + image_acc1d_write.use(); + }); + }); - accessor + sycl::accessor image_acc2d_write; - kernel( - [=]() { - image_acc2d_write.use(); - }); + q.submit([&](sycl::handler &h) { + h.single_task( + [=] { + image_acc2d_write.use(); + }); + }); - accessor + sycl::accessor image_acc3d_write; - kernel( - [=]() { - image_acc3d_write.use(); - }); + q.submit([&](sycl::handler &h) { + h.single_task( + [=] { + image_acc3d_write.use(); + }); + }); } // CHECK: {{.*}}use_image1d_r 'void (__read_only image1d_t)' diff --git a/clang/test/SemaSYCL/accessors-targets.cpp b/clang/test/SemaSYCL/accessors-targets.cpp index 60b5b9159fc14..9b05d2ce10298 100644 --- a/clang/test/SemaSYCL/accessors-targets.cpp +++ b/clang/test/SemaSYCL/accessors-targets.cpp @@ -1,41 +1,47 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s -// This test checks that compiler generates correct kernel wrapper arguments for +// This test checks that the compiler generates correct kernel wrapper arguments for // different accessors targets. -#include "Inputs/sycl.hpp" +#include "sycl.hpp" -using namespace cl::sycl; - -template -__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) { - kernelFunc(); -} +sycl::queue q; int main() { - - accessor + // Access work-group local memory with read and write access. + sycl::accessor local_acc; - accessor + // Access buffer via global memory with read and write access. + sycl::accessor global_acc; - accessor + // Access buffer via constant memory with read and write access. + sycl::accessor constant_acc; - kernel( - [=]() { - local_acc.use(); - }); - kernel( - [=]() { - global_acc.use(); - }); - kernel( - [=]() { - constant_acc.use(); - }); + + q.submit([&](sycl::handler &h) { + h.single_task( + [=] { + local_acc.use(); + }); + }); + + q.submit([&](sycl::handler &h) { + h.single_task( + [=] { + global_acc.use(); + }); + }); + + q.submit([&](sycl::handler &h) { + h.single_task( + [=] { + constant_acc.use(); + }); + }); } -// CHECK: {{.*}}use_local{{.*}} 'void (__local int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>)' -// CHECK: {{.*}}use_global{{.*}} 'void (__global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>)' -// CHECK: {{.*}}use_constant{{.*}} 'void (__constant int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>)' +// CHECK: {{.*}}use_local{{.*}} 'void (__local int *, sycl::range<1>, sycl::range<1>, sycl::id<1>)' +// CHECK: {{.*}}use_global{{.*}} 'void (__global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>)' +// CHECK: {{.*}}use_constant{{.*}} 'void (__constant int *, sycl::range<1>, sycl::range<1>, sycl::id<1>)' diff --git a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp index 6b17af6351e22..6ab66376291b2 100644 --- a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp +++ b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp @@ -1,9 +1,10 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -fcxx-exceptions -Wno-return-type -Wno-sycl-2017-compat -verify -fsyntax-only -std=c++20 -Werror=vla %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -fcxx-exceptions -Wno-return-type -sycl-std=2020 -verify -fsyntax-only -std=c++20 -Werror=vla %s -template -__attribute__((sycl_kernel)) void kernel_single_task(const Func &kernelFunc) { - kernelFunc(); -} +// This test verifies that a SYCL kernel executed on a device, cannot call a recursive function. + +#include "sycl.hpp" + +sycl::queue q; // expected-note@+1{{function implemented using recursion declared here}} constexpr int constexpr_recurse1(int n); @@ -71,6 +72,11 @@ void constexpr_recurse_test_err() { } int main() { - kernel_single_task([]() { constexpr_recurse_test(); }); - kernel_single_task([]() { constexpr_recurse_test_err(); }); + q.submit([&](sycl::handler &h) { + h.single_task([]() { constexpr_recurse_test(); }); + }); + + q.submit([&](sycl::handler &h) { + h.single_task([]() { constexpr_recurse_test_err(); }); + }); } diff --git a/clang/test/SemaSYCL/array-kernel-param-neg.cpp b/clang/test/SemaSYCL/array-kernel-param-neg.cpp index 5ad8569feebe8..b0a5bade19e84 100755 --- a/clang/test/SemaSYCL/array-kernel-param-neg.cpp +++ b/clang/test/SemaSYCL/array-kernel-param-neg.cpp @@ -1,43 +1,49 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -fcxx-exceptions -Wno-sycl-2017-compat -verify -fsyntax-only %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -fcxx-exceptions -sycl-std=2020 -verify -fsyntax-only %s // This test checks if compiler reports compilation error on an attempt to pass // an array of non-trivially copyable structs as SYCL kernel parameter or // a non-constant size array. -struct B { +#include "sycl.hpp" + +sycl::queue q; + +struct NonTrivialCopyStruct { int i; - B(int _i) : i(_i) {} - B(const B &x) : i(x.i) {} + NonTrivialCopyStruct(int _i) : i(_i) {} + NonTrivialCopyStruct(const NonTrivialCopyStruct &x) : i(x.i) {} }; -struct D { +struct NonTrivialDestructorStruct { int i; - ~D(); + ~NonTrivialDestructorStruct(); }; -class E { +class Array { // expected-error@+1 {{kernel parameter is not a constant size array}} - int i[]; + int NonConstantSizeArray[]; public: - int operator()() const { return i[0]; } + int operator()() const { return NonConstantSizeArray[0]; } }; -template -__attribute__((sycl_kernel)) void kernel_single_task(const Func &kernelFunc) { - kernelFunc(); -} - void test() { - B nsl1[4] = {1, 2, 3, 4}; - D nsl2[5]; - E es; - kernel_single_task([=] { - // expected-error@+1 {{kernel parameter has non-trivially copy constructible class/struct type}} - int b = nsl1[2].i; - // expected-error@+1 {{kernel parameter has non-trivially destructible class/struct type}} - int d = nsl2[4].i; + NonTrivialCopyStruct NTCSObject[4] = {1, 2, 3, 4}; + NonTrivialDestructorStruct NTDSObject[5]; + // expected-note@+1 {{'UnknownSizeArrayObj' declared here}} + Array UnknownSizeArrayObj; + + q.submit([&](sycl::handler &h) { + h.single_task([=] { + // expected-error@+1 {{kernel parameter has non-trivially copy constructible class/struct type}} + int b = NTCSObject[2].i; + // expected-error@+1 {{kernel parameter has non-trivially destructible class/struct type}} + int d = NTDSObject[4].i; + }); }); - kernel_single_task(es); + q.submit([&](sycl::handler &h) { + // expected-error@+1 {{variable 'UnknownSizeArrayObj' with flexible array member cannot be captured in a lambda expression}} + h.single_task(UnknownSizeArrayObj); + }); } diff --git a/clang/test/SemaSYCL/array-kernel-param.cpp b/clang/test/SemaSYCL/array-kernel-param.cpp index 58eee1c4debef..9701e331ad236 100644 --- a/clang/test/SemaSYCL/array-kernel-param.cpp +++ b/clang/test/SemaSYCL/array-kernel-param.cpp @@ -1,16 +1,13 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s // This test checks that compiler generates correct kernel arguments for // arrays, Accessor arrays, and structs containing Accessors. -#include "Inputs/sycl.hpp" +#include "sycl.hpp" -using namespace cl::sycl; +sycl::queue myQueue; -template -__attribute__((sycl_kernel)) void a_kernel(const Func &kernelFunc) { - kernelFunc(); -} +using namespace sycl; template struct S { @@ -20,102 +17,119 @@ struct S { int main() { using Accessor = - accessor; + sycl::accessor; - Accessor acc[2]; - int a[2]; - int *a_ptrs[2]; + Accessor ReadWriteAccessor[2]; + int Array[2]; + int *ArrayOfPointers[2]; - struct struct_acc_t { + struct StructWithAccessors { Accessor member_acc[2]; - } struct_acc; + } StructAccArrayObj; + S s; - struct foo_inner { - int foo_inner_x; - int foo_inner_y; - int *foo_inner_z[2]; + struct StructWithPointers { + int x; + int y; + int *ArrayOfPtrs[2]; }; - struct foo { - int foo_a; - foo_inner foo_b[2]; - int *foo_2D[2][1]; - int foo_c; + struct DecomposedStruct { + int a; + StructWithPointers SWPtrsMem[2]; + int *Array_2D_Ptrs[2][1]; + int c; }; // Not decomposed. - struct foo2 { - int foo_a; - int foo_2D[2][1]; - int foo_c; + struct NonDecomposedStruct { + int a; + int Array_2D[2][1]; + int c; }; - foo struct_array[2]; - foo2 struct_array2[2]; + DecomposedStruct DecompStructArray[2]; + NonDecomposedStruct NonDecompStructArray[2]; int array_2D[2][3]; - a_kernel( - [=]() { - acc[1].use(); - }); - - a_kernel( - [=]() { - int local = a[1]; - }); - - a_kernel( - [=]() { - int local = *a_ptrs[1]; - }); - - a_kernel( - [=]() { - struct_acc.member_acc[2].use(); - }); - - a_kernel( - [=]() { - foo local = struct_array[1]; - }); - - a_kernel( - [=]() { - int local = s.a[2]; - }); - - a_kernel( - [=]() { - int local = array_2D[1][1]; - }); - - a_kernel( - [=]() { - foo2 local = struct_array2[0]; - }); + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + ReadWriteAccessor[1].use(); + }); + }); + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + int local = Array[1]; + }); + }); + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + int local = *ArrayOfPointers[1]; + }); + }); + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + StructAccArrayObj.member_acc[2].use(); + }); + }); + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + DecomposedStruct local = DecompStructArray[1]; + }); + }); + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + int local = s.a[2]; + }); + }); + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + int local = array_2D[1][1]; + }); + }); + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + NonDecomposedStruct local = NonDecompStructArray[0]; + }); + }); } -// Check kernel_A parameters -// CHECK: FunctionDecl {{.*}}kernel_A{{.*}} 'void (__global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, __global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>)' +// Check Kernel_Accessor parameters +// CHECK: FunctionDecl {{.*}}Kernel_Accessor{{.*}} 'void (__global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, __global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>)' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ '__global int *' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'cl::sycl::range<1>' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'cl::sycl::range<1>' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'cl::sycl::id<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'sycl::range<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'sycl::range<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'sycl::id<1>' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ '__global int *' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'cl::sycl::range<1>' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'cl::sycl::range<1>' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'cl::sycl::id<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'sycl::range<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'sycl::range<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'sycl::id<1>' // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}}__init // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}}__init -// Check kernel_B parameters -// CHECK: FunctionDecl {{.*}}kernel_B{{.*}} 'void (__wrapper_class)' +// Check Kernel_Array parameters +// CHECK: FunctionDecl {{.*}}Kernel_Array{{.*}} 'void (__wrapper_class)' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ '__wrapper_class' -// Check kernel_B inits +// Check Kernel_Array inits // CHECK-NEXT: CompoundStmt // CHECK-NEXT: DeclStmt // CHECK-NEXT: VarDecl {{.*}} cinit @@ -131,11 +145,11 @@ int main() { // CHECK-NEXT: MemberExpr {{.*}} 'int [2]' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_' '__wrapper_class' -// Check kernel_B_ptrs parameters -// CHECK: FunctionDecl {{.*}}kernel_B_ptrs{{.*}} 'void (__global int *, __global int *)' +// Check Kernel_Array_Ptrs parameters +// CHECK: FunctionDecl {{.*}}Kernel_Array_Ptrs{{.*}} 'void (__global int *, __global int *)' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ '__global int *' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ '__global int *' -// Check kernel_B_ptrs inits +// Check Kernel_Array_Ptrs inits // CHECK-NEXT: CompoundStmt // CHECK-NEXT: DeclStmt // CHECK-NEXT: VarDecl {{.*}} cinit @@ -148,21 +162,21 @@ int main() { // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: DeclRefExpr {{.*}} '__global int *' lvalue ParmVar {{.*}} '_arg_' '__global int *' -// Check kernel_C parameters -// CHECK: FunctionDecl {{.*}}kernel_C{{.*}} 'void (__global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, __global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>)' +// Check Kernel_StructAccArray parameters +// CHECK: FunctionDecl {{.*}}Kernel_StructAccArray{{.*}} 'void (__global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, __global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>)' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc '__global int *' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'cl::sycl::range<1>' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'cl::sycl::range<1>' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'cl::sycl::id<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'sycl::range<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'sycl::range<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'sycl::id<1>' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc '__global int *' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'cl::sycl::range<1>' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'cl::sycl::range<1>' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'cl::sycl::id<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'sycl::range<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'sycl::range<1>' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_member_acc 'sycl::id<1>' // CHECK-NEXT: CompoundStmt // CHECK-NEXT: DeclStmt // CHECK-NEXT: VarDecl {{.*}} used '(lambda at {{.*}}array-kernel-param.cpp{{.*}})' cinit // CHECK-NEXT: InitListExpr {{.*}} '(lambda at {{.*}}array-kernel-param.cpp{{.*}})' -// CHECK-NEXT: InitListExpr {{.*}} 'struct_acc_t' +// CHECK-NEXT: InitListExpr {{.*}} 'StructWithAccessors' // CHECK-NEXT: InitListExpr {{.*}} 'Accessor [2]' // CHECK-NEXT: CXXConstructExpr {{.*}} 'Accessor' // CHECK-NEXT: CXXConstructExpr {{.*}} 'Accessor' @@ -173,142 +187,142 @@ int main() { // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}}__init -// Check kernel_D parameters -// CHECK: FunctionDecl {{.*}}kernel_D{{.*}} 'void (int, int, int, __wrapper_class, __wrapper_class, int, int, __wrapper_class, __wrapper_class, __wrapper_class, __wrapper_class, int, int, int, int, __wrapper_class, __wrapper_class, int, int, __wrapper_class, __wrapper_class, __wrapper_class, __wrapper_class, int)' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_a 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_x 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_y 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_z '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_z '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_x 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_y 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_z '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_z '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_2D '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_2D '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_c 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_a 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_x 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_y 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_z '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_z '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_x 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_y 'int' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_z '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_inner_z '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_2D '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_2D '__wrapper_class' -// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_foo_c 'int' +// Check Kernel_DecomposedStruct parameters +// CHECK: FunctionDecl {{.*}}Kernel_DecomposedStruct{{.*}} 'void (int, int, int, __wrapper_class, __wrapper_class, int, int, __wrapper_class, __wrapper_class, __wrapper_class, __wrapper_class, int, int, int, int, __wrapper_class, __wrapper_class, int, int, __wrapper_class, __wrapper_class, __wrapper_class, __wrapper_class, int)' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_a 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_x 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_y 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ArrayOfPtrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ArrayOfPtrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_x 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_y 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ArrayOfPtrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ArrayOfPtrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_Array_2D_Ptrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_Array_2D_Ptrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_c 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_a 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_x 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_y 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ArrayOfPtrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ArrayOfPtrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_x 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_y 'int' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ArrayOfPtrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ArrayOfPtrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_Array_2D_Ptrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_Array_2D_Ptrs '__wrapper_class' +// CHECK-NEXT: ParmVarDecl {{.*}} used _arg_c 'int' // CHECK-NEXT: CompoundStmt // CHECK-NEXT: DeclStmt // CHECK-NEXT: VarDecl {{.*}} used '(lambda at {{.*}}array-kernel-param.cpp{{.*}})' cinit // CHECK-NEXT: InitListExpr {{.*}} '(lambda at {{.*}}array-kernel-param.cpp{{.*}})' -// Initializer for struct array i.e. foo struct_array[2] -// CHECK-NEXT: InitListExpr {{.*}} 'foo [2]' +// Initializer for struct array i.e. DecomposedStruct DecompStructArray[2] +// CHECK-NEXT: InitListExpr {{.*}} 'DecomposedStruct [2]' -// Initializer for first element of struct_array -// CHECK-NEXT: InitListExpr {{.*}} 'foo' +// Initializer for first element of DecompStructArray +// CHECK-NEXT: InitListExpr {{.*}} 'DecomposedStruct' // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_a' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_a' 'int' -// Initializer for struct array inside foo i.e. foo_inner foo_b[2] -// CHECK-NEXT: InitListExpr {{.*}} 'foo_inner [2]' +// Initializer for struct array inside DecomposedStruct i.e. StructWithPointers SWPtrsMem[2] +// CHECK-NEXT: InitListExpr {{.*}} 'StructWithPointers [2]' // Initializer for first element of inner struct array -// CHECK-NEXT: InitListExpr {{.*}} 'foo_inner' +// CHECK-NEXT: InitListExpr {{.*}} 'StructWithPointers' // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_x' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_x' 'int' // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_y' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_y' 'int' // CHECK-NEXT: InitListExpr {{.*}} 'int *[2]' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_inner_z' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_ArrayOfPtrs' '__wrapper_class' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_inner_z' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_ArrayOfPtrs' '__wrapper_class' // Initializer for second element of inner struct array -// CHECK-NEXT: InitListExpr {{.*}} 'foo_inner' +// CHECK-NEXT: InitListExpr {{.*}} 'StructWithPointers' // CHECK-NEXT: ImplicitCastExpr -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_x' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_x' 'int' // CHECK-NEXT: ImplicitCastExpr -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_y' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_y' 'int' // CHECK-NEXT: InitListExpr {{.*}} 'int *[2]' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_inner_z' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_ArrayOfPtrs' '__wrapper_class' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_inner_z' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_ArrayOfPtrs' '__wrapper_class' // CHECK-NEXT: InitListExpr {{.*}} 'int *[2][1]' // CHECK-NEXT: InitListExpr {{.*}} 'int *[1]' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_2D' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_Array_2D_Ptrs' '__wrapper_class' // CHECK-NEXT: InitListExpr {{.*}} 'int *[1]' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_2D' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_Array_2D_Ptrs' '__wrapper_class' // CHECK-NEXT: ImplicitCastExpr -// CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_c' 'int' +// CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar {{.*}} '_arg_c' 'int' -// Initializer for second element of struct_array -// CHECK-NEXT: InitListExpr {{.*}} 'foo' +// Initializer for second element of DecompStructArray +// CHECK-NEXT: InitListExpr {{.*}} 'DecomposedStruct' // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_a' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_a' 'int' -// Initializer for struct array inside foo i.e. foo_inner foo_b[2] -// CHECK-NEXT: InitListExpr {{.*}} 'foo_inner [2]' +// Initializer for struct array inside DecomposedStruct i.e. StructWithPointers SWPtrsMem[2] +// CHECK-NEXT: InitListExpr {{.*}} 'StructWithPointers [2]' // Initializer for first element of inner struct array -// CHECK-NEXT: InitListExpr {{.*}} 'foo_inner' +// CHECK-NEXT: InitListExpr {{.*}} 'StructWithPointers' // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_x' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_x' 'int' // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_y' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_y' 'int' // CHECK-NEXT: InitListExpr {{.*}} 'int *[2]' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_inner_z' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_ArrayOfPtrs' '__wrapper_class' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_inner_z' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_ArrayOfPtrs' '__wrapper_class' // Initializer for second element of inner struct array -// CHECK-NEXT: InitListExpr {{.*}} 'foo_inner' +// CHECK-NEXT: InitListExpr {{.*}} 'StructWithPointers' // CHECK-NEXT: ImplicitCastExpr -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_x' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_x' 'int' // CHECK-NEXT: ImplicitCastExpr -// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_y' 'int' +// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_y' 'int' // CHECK-NEXT: InitListExpr {{.*}} 'int *[2]' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_inner_z' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_ArrayOfPtrs' '__wrapper_class' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_inner_z' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_ArrayOfPtrs' '__wrapper_class' // CHECK-NEXT: InitListExpr {{.*}} 'int *[2][1]' // CHECK-NEXT: InitListExpr {{.*}} 'int *[1]' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_2D' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_Array_2D_Ptrs' '__wrapper_class' // CHECK-NEXT: InitListExpr {{.*}} 'int *[1]' // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global int *' // CHECK-NEXT: MemberExpr {{.*}} '__global int *' lvalue . -// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_foo_2D' '__wrapper_class' +// CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_Array_2D_Ptrs' '__wrapper_class' -// Check kernel_E parameters -// CHECK: FunctionDecl {{.*}}kernel_E{{.*}} 'void (S)' +// Check Kernel_TemplatedStructArray parameters +// CHECK: FunctionDecl {{.*}}Kernel_TemplatedStructArray{{.*}} 'void (S)' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'S':'S' // CHECK-NEXT: CompoundStmt // CHECK-NEXT: DeclStmt @@ -318,10 +332,10 @@ int main() { // CHECK-NEXT: ImplicitCastExpr // CHECK-NEXT: DeclRefExpr {{.*}} 'S':'S' lvalue ParmVar {{.*}} '_arg_' 'S':'S' -// Check kernel_F parameters -// CHECK: FunctionDecl {{.*}}kernel_F{{.*}} 'void (__wrapper_class)' +// Check Kernel_Array_2D parameters +// CHECK: FunctionDecl {{.*}}Kernel_Array_2D{{.*}} 'void (__wrapper_class)' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ '__wrapper_class' -// Check kernel_F inits +// Check Kernel_Array_2D inits // CHECK-NEXT: CompoundStmt // CHECK-NEXT: DeclStmt // CHECK-NEXT: VarDecl {{.*}} cinit @@ -350,23 +364,23 @@ int main() { // CHECK-NEXT: ArrayInitIndexExpr {{.*}} 'unsigned // CHECK-NEXT: ArrayInitIndexExpr {{.*}} 'unsigned -// Check kernel_G parameters. -// CHECK: FunctionDecl {{.*}}kernel_G{{.*}} 'void (__wrapper_class)' +// Check Kernel_NonDecomposedStruct parameters. +// CHECK: FunctionDecl {{.*}}Kernel_NonDecomposedStruct{{.*}} 'void (__wrapper_class)' // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ '__wrapper_class' -// Check kernel_G inits +// Check Kernel_NonDecomposedStruct inits // CHECK-NEXT: CompoundStmt // CHECK-NEXT: DeclStmt // CHECK-NEXT: VarDecl {{.*}} cinit // CHECK-NEXT: InitListExpr -// CHECK-NEXT: ArrayInitLoopExpr {{.*}} 'foo2 [2]' -// CHECK-NEXT: OpaqueValueExpr {{.*}} 'foo2 [2]' lvalue -// CHECK-NEXT: MemberExpr {{.*}} 'foo2 [2]' lvalue . +// CHECK-NEXT: ArrayInitLoopExpr {{.*}} 'NonDecomposedStruct [2]' +// CHECK-NEXT: OpaqueValueExpr {{.*}} 'NonDecomposedStruct [2]' lvalue +// CHECK-NEXT: MemberExpr {{.*}} 'NonDecomposedStruct [2]' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_' '__wrapper_class' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'foo2' 'void (const foo2 &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const foo2' lvalue -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'foo2' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'foo2 *' -// CHECK-NEXT: OpaqueValueExpr {{.*}} 'foo2 [2]' lvalue -// CHECK-NEXT: MemberExpr {{.*}} 'foo2 [2]' lvalue . +// CHECK-NEXT: CXXConstructExpr {{.*}} 'NonDecomposedStruct' 'void (const NonDecomposedStruct &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const NonDecomposedStruct' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'NonDecomposedStruct' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'NonDecomposedStruct *' +// CHECK-NEXT: OpaqueValueExpr {{.*}} 'NonDecomposedStruct [2]' lvalue +// CHECK-NEXT: MemberExpr {{.*}} 'NonDecomposedStruct [2]' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '__wrapper_class' lvalue ParmVar {{.*}} '_arg_' '__wrapper_class' // CHECK-NEXT: ArrayInitIndexExpr {{.*}} 'unsigned diff --git a/clang/test/SemaSYCL/basic-kernel-wrapper.cpp b/clang/test/SemaSYCL/basic-kernel-wrapper.cpp index f1ff38ec993df..6176fa3098322 100644 --- a/clang/test/SemaSYCL/basic-kernel-wrapper.cpp +++ b/clang/test/SemaSYCL/basic-kernel-wrapper.cpp @@ -1,36 +1,33 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s // This test checks that compiler generates correct kernel wrapper for basic // case. -#include "Inputs/sycl.hpp" +#include "sycl.hpp" -template -struct AccWrapper { Acc accessor; }; - -template -__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) { - kernelFunc(); -} +sycl::queue myQueue; int main() { - cl::sycl::accessor acc; - kernel( - [=]() { - acc.use(); - }); + sycl::accessor readWriteAccessor; + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=]() { + readWriteAccessor.use(); + }); + }); } // Check declaration of the kernel -// CHECK: FunctionDecl {{.*}}kernel_wrapper{{.*}} 'void (__global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>)' +// CHECK: FunctionDecl {{.*}}kernel_wrapper{{.*}} 'void (__global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>)' // Check parameters of the kernel // CHECK: ParmVarDecl {{.*}} used [[_arg_Mem:[0-9a-zA-Z_]+]] '__global int *' -// CHECK: ParmVarDecl {{.*}} used [[_arg_AccessRange:[0-9a-zA-Z_]+]] 'cl::sycl::range<1>' -// CHECK: ParmVarDecl {{.*}} used [[_arg_MemRange:[0-9a-zA-Z_]+]] 'cl::sycl::range<1>' -// CHECK: ParmVarDecl {{.*}} used [[_arg_Offset:[0-9a-zA-Z_]+]] 'cl::sycl::id<1>' +// CHECK: ParmVarDecl {{.*}} used [[_arg_AccessRange:[0-9a-zA-Z_]+]] 'sycl::range<1>' +// CHECK: ParmVarDecl {{.*}} used [[_arg_MemRange:[0-9a-zA-Z_]+]] 'sycl::range<1>' +// CHECK: ParmVarDecl {{.*}} used [[_arg_Offset:[0-9a-zA-Z_]+]] 'sycl::id<1>' // Check body of the kernel @@ -43,23 +40,23 @@ int main() { // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ({{.*}}PtrType, range<1>, range<1>, id<1>)' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::accessor':'cl::sycl::accessor' lvalue . +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::accessor':'sycl::accessor' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at {{.*}}basic-kernel-wrapper.cpp{{.*}})' lvalue Var // CHECK-NEXT: ImplicitCastExpr {{.*}} // CHECK-NEXT: DeclRefExpr {{.*}} '__global int *' lvalue ParmVar {{.*}} '[[_arg_Mem]]' '__global int *' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'range<1>':'cl::sycl::range<1>' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::range<1>' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::range<1>' lvalue ParmVar {{.*}} '[[_arg_AccessRange]]' 'cl::sycl::range<1>' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'range<1>':'sycl::range<1>' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::range<1>' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::range<1>' lvalue ParmVar {{.*}} '[[_arg_AccessRange]]' 'sycl::range<1>' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'range<1>':'cl::sycl::range<1>' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::range<1>' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::range<1>' lvalue ParmVar {{.*}} '[[_arg_MemRange]]' 'cl::sycl::range<1>' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'range<1>':'sycl::range<1>' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::range<1>' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::range<1>' lvalue ParmVar {{.*}} '[[_arg_MemRange]]' 'sycl::range<1>' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'id<1>':'cl::sycl::id<1>' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::id<1>' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::id<1>' lvalue ParmVar {{.*}} '[[_arg_Offset]]' 'cl::sycl::id<1>' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'id<1>':'sycl::id<1>' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::id<1>' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::id<1>' lvalue ParmVar {{.*}} '[[_arg_Offset]]' 'sycl::id<1>' // Check that body of the kernel caller function is included into kernel diff --git a/clang/test/SemaSYCL/decomposition.cpp b/clang/test/SemaSYCL/decomposition.cpp index 6830c944f239c..83ac5a6cdc653 100644 --- a/clang/test/SemaSYCL/decomposition.cpp +++ b/clang/test/SemaSYCL/decomposition.cpp @@ -1,124 +1,156 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s -#include "Inputs/sycl.hpp" +// This test checks that the compiler decomposes structs containing special types only +// (i.e. accessor/stream/sampler etc) and all others are passed without decomposition +// thus optimizing the number of kernel arguments. -using namespace cl::sycl; +#include "sycl.hpp" -struct has_acc { - accessor acc; +sycl::queue myQueue; + +struct StructWithAccessor { + sycl::accessor acc; }; -struct acc_base : accessor { +struct StructInheritedAccessor : sycl::accessor { int i; }; -struct has_sampler { - sampler sampl; +struct StructWithSampler { + sycl::sampler sampl; }; -struct has_spec_const { - ONEAPI::experimental::spec_constant SC; +struct StructWithSpecConst { + sycl::ONEAPI::experimental::spec_constant SC; }; -handler H; +sycl::handler H; -struct has_stream { - stream s1{0, 0, H}; +struct StructWithStream { + sycl::stream s1{0, 0, H}; }; -struct has_half { - half h; +struct StructWithHalf { + sycl::half h; }; -struct non_decomposed { +struct StructNonDecomposed { int i; float f; double d; }; -struct use_non_decomposed : non_decomposed { - non_decomposed member; +struct StructWithNonDecomposedStruct : StructNonDecomposed { + StructNonDecomposed member; float f; double d; }; template -struct Test1 { +struct StructWithArray { T a; T b[2]; - non_decomposed d; + StructNonDecomposed SimpleStruct; int i; }; template -struct Test2 : T { - non_decomposed d; +struct DerivedStruct : T { + StructNonDecomposed SimpleStruct; int i; }; -template -__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) { - kernelFunc(); -} - int main() { - non_decomposed d; - non_decomposed ds[5]; - use_non_decomposed d2; - use_non_decomposed d2s[5]; + StructNonDecomposed SimpleStruct; + StructNonDecomposed ArrayOfSimpleStruct[5]; + StructWithNonDecomposedStruct NonDecompStruct; + StructWithNonDecomposedStruct ArrayOfNonDecompStruct[5]; // Check to ensure that these are not decomposed. - kernel([=]() { return d.i + ds[0].i + d2.i + d2s[0].i; }); - // CHECK: FunctionDecl {{.*}}NonDecomp{{.*}} 'void (non_decomposed, __wrapper_class, use_non_decomposed, __wrapper_class)' + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return SimpleStruct.i + ArrayOfSimpleStruct[0].i + NonDecompStruct.i + ArrayOfNonDecompStruct[0].i; }); + }); + // CHECK: FunctionDecl {{.*}}NonDecomposed{{.*}} 'void (StructNonDecomposed, __wrapper_class, StructWithNonDecomposedStruct, __wrapper_class)' { - Test1 t1; - kernel([=]() { return t1.i; }); - // CHECK: FunctionDecl {{.*}}Acc1{{.*}} 'void (__global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, __global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, __global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, non_decomposed, int)' - Test2 t2; - kernel([=]() { return t2.i; }); - // CHECK: FunctionDecl {{.*}}Acc2{{.*}} 'void (__global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, non_decomposed, int)' - Test1 t3; - kernel([=]() { return t3.i; }); - // CHECK: FunctionDecl {{.*}}Acc3{{.*}} 'void (__global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, int, __global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, int, __global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, int, non_decomposed, int)' - Test2 t4; - kernel([=]() { return t4.i; }); - // CHECK: FunctionDecl {{.*}}Acc4{{.*}} 'void (__global char *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, int, non_decomposed, int)' + StructWithArray t1; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t1.i; }); + }); + // CHECK: FunctionDecl {{.*}}Acc1{{.*}} 'void (__global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, __global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, __global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, StructNonDecomposed, int)' + + DerivedStruct t2; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t2.i; }); + }); + // CHECK: FunctionDecl {{.*}}Acc2{{.*}} 'void (__global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, StructNonDecomposed, int)' + + StructWithArray t3; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t3.i; }); + }); + // CHECK: FunctionDecl {{.*}}Acc3{{.*}} 'void (__global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, int, __global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, int, __global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, int, StructNonDecomposed, int)' + + DerivedStruct t4; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t4.i; }); + }); + // CHECK: FunctionDecl {{.*}}Acc4{{.*}} 'void (__global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, int, StructNonDecomposed, int)' } { - Test1 t1; - kernel([=]() { return t1.i; }); - // CHECK: FunctionDecl {{.*}}Sampl1{{.*}} 'void (sampler_t, sampler_t, sampler_t, non_decomposed, int)' - Test2 t2; - kernel([=]() { return t2.i; }); - // CHECK: FunctionDecl {{.*}}Sampl2{{.*}} 'void (sampler_t, non_decomposed, int)' + StructWithArray t1; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t1.i; }); + }); + // CHECK: FunctionDecl {{.*}}Sampl1{{.*}} 'void (sampler_t, sampler_t, sampler_t, StructNonDecomposed, int)' + + DerivedStruct t2; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t2.i; }); + }); + // CHECK: FunctionDecl {{.*}}Sampl2{{.*}} 'void (sampler_t, StructNonDecomposed, int)' } { - Test1 t1; - kernel([=]() { return t1.i; }); - // CHECK: FunctionDecl {{.*}}SpecConst{{.*}} 'void (non_decomposed, int)' - Test2 t2; - kernel([=]() { return t2.i; }); - // CHECK: FunctionDecl {{.*}}SpecConst2{{.*}} 'void (non_decomposed, int)' + StructWithArray t1; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t1.i; }); + }); + // CHECK: FunctionDecl {{.*}}SpecConst{{.*}} 'void (StructNonDecomposed, int)' + + DerivedStruct t2; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t2.i; }); + }); + // CHECK: FunctionDecl {{.*}}SpecConst2{{.*}} 'void (StructNonDecomposed, int)' } { - Test1 t1; - kernel([=]() { return t1.i; }); - // CHECK: FunctionDecl {{.*}}Stream1{{.*}} 'void (cl::sycl::stream, __global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, cl::sycl::stream, __global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, cl::sycl::stream, __global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, non_decomposed, int)' - Test2 t2; - kernel([=]() { return t2.i; }); - // CHECK: FunctionDecl {{.*}}Stream2{{.*}} 'void (cl::sycl::stream, __global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, non_decomposed, int)' + StructWithArray t1; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t1.i; }); + }); + // CHECK: FunctionDecl {{.*}}Stream1{{.*}} 'void (sycl::stream, __global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, sycl::stream, __global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, sycl::stream, __global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, StructNonDecomposed, int)' + + DerivedStruct t2; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t2.i; }); + }); + // CHECK: FunctionDecl {{.*}}Stream2{{.*}} 'void (sycl::stream, __global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, StructNonDecomposed, int)' } { - Test1 t1; - kernel([=]() { return t1.i; }); - // CHECK: FunctionDecl {{.*}}Half1{{.*}} 'void (cl::sycl::half, cl::sycl::half, cl::sycl::half, non_decomposed, int)' - Test2 t2; - kernel([=]() { return t2.i; }); - // CHECK: FunctionDecl {{.*}}Half2{{.*}} 'void (cl::sycl::half, non_decomposed, int)' + StructWithArray t1; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t1.i; }); + }); + // CHECK: FunctionDecl {{.*}}Half1{{.*}} 'void (sycl::half, sycl::half, sycl::half, StructNonDecomposed, int)' + + DerivedStruct t2; + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { return t2.i; }); + }); + // CHECK: FunctionDecl {{.*}}Half2{{.*}} 'void (sycl::half, StructNonDecomposed, int)' } } diff --git a/clang/test/SemaSYCL/fake-accessors.cpp b/clang/test/SemaSYCL/fake-accessors.cpp index d4dd2a0f60209..374eb320b71ff 100644 --- a/clang/test/SemaSYCL/fake-accessors.cpp +++ b/clang/test/SemaSYCL/fake-accessors.cpp @@ -1,8 +1,10 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -Wno-int-to-void-pointer-cast -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s -#include "Inputs/sycl.hpp" +#include "sycl.hpp" -namespace foo { +sycl::queue deviceQueue; + +namespace fake { namespace cl { namespace sycl { class accessor { @@ -11,46 +13,46 @@ class accessor { }; } // namespace sycl } // namespace cl -} // namespace foo +} // namespace fake class accessor { public: int field; }; -typedef cl::sycl::accessor - MyAccessorTD; +int main() { -using MyAccessorA = cl::sycl::accessor; + fake::cl::sycl::accessor FakeAccessor = {1}; + accessor AccessorClass = {1}; -template -__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) { - kernelFunc(); -} + typedef sycl::accessor + MyAccessorTD; + MyAccessorTD AccessorTypeDef; -int main() { - foo::cl::sycl::accessor acc = {1}; - accessor acc1 = {1}; - - cl::sycl::accessor accessorA; - cl::sycl::accessor accessorB; - cl::sycl::accessor accessorC; - kernel( - [=]() { - accessorA.use((void*)(acc.field + acc1.field)); + using MyAccessorA = sycl::accessor; + MyAccessorA AccessorAlias; + + cl::sycl::accessor AccessorRegular; + + deviceQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + AccessorRegular.use((void *)(FakeAccessor.field + AccessorClass.field)); }); - kernel( - [=]() { - accessorB.use((void*)(acc.field + acc1.field)); + + h.single_task( + [=] { + AccessorTypeDef.use((void *)(FakeAccessor.field + AccessorClass.field)); }); - kernel( - [=]() { - accessorC.use((void*)(acc.field + acc1.field)); + + h.single_task( + [=] { + AccessorAlias.use((void *)(FakeAccessor.field + AccessorClass.field)); }); + }); + return 0; } -// CHECK: fake_accessors{{.*}} 'void (__global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, foo::cl::sycl::accessor, accessor) -// CHECK: accessor_typedef{{.*}} 'void (__global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, foo::cl::sycl::accessor, accessor) -// CHECK: accessor_alias{{.*}} 'void (__global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>, foo::cl::sycl::accessor, accessor) +// CHECK: fake_accessors{{.*}} 'void (__global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, fake::cl::sycl::accessor, accessor) +// CHECK: accessor_typedef{{.*}} 'void (__global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, fake::cl::sycl::accessor, accessor) +// CHECK: accessor_alias{{.*}} 'void (__global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, fake::cl::sycl::accessor, accessor) diff --git a/clang/test/SemaSYCL/half-kernel-arg.cpp b/clang/test/SemaSYCL/half-kernel-arg.cpp index e6297351fad8b..77bbafc3eb2bc 100644 --- a/clang/test/SemaSYCL/half-kernel-arg.cpp +++ b/clang/test/SemaSYCL/half-kernel-arg.cpp @@ -1,23 +1,27 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s // This test checks that compiler generates correct initialization for arguments -// that have cl::sycl::half type inside the OpenCL kernel +// that have sycl::half type inside the OpenCL kernel -#include "Inputs/sycl.hpp" +#include "sycl.hpp" + +sycl::queue myQueue; int main() { - cl::sycl::half HostHalf; - cl::sycl::kernel_single_task( - [=]() { - cl::sycl::half KernelHalf = HostHalf; - }); + sycl::half HostHalf; + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=]() { + sycl::half KernelHalf = HostHalf; + }); + }); } -// CHECK: {{.*}}kernel_half{{.*}} 'void (cl::sycl::half)' -// CHECK: ParmVarDecl {{.*}} used _arg_ 'cl::sycl::half':'cl::sycl::detail::half_impl::half' +// CHECK: {{.*}}kernel_half{{.*}} 'void (sycl::half)' +// CHECK: ParmVarDecl {{.*}} used _arg_ 'sycl::half':'sycl::detail::half_impl::half' // // Check that lambda field of half type is initialized // CHECK: VarDecl {{.*}}'(lambda at {{.*}}' // CHECK-NEXT: InitListExpr {{.*}}'(lambda at {{.*}}' -// CHECK-NEXT: CXXConstructExpr {{.*}}'cl::sycl::detail::half_impl::half'{{.*}} -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::detail::half_impl::half' -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::half':'cl::sycl::detail::half_impl::half' lvalue ParmVar {{.*}} '_arg_' 'cl::sycl::half':'cl::sycl::detail::half_impl::half' +// CHECK-NEXT: CXXConstructExpr {{.*}}'sycl::detail::half_impl::half'{{.*}} +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::detail::half_impl::half' +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::half':'sycl::detail::half_impl::half' lvalue ParmVar {{.*}} '_arg_' 'sycl::half':'sycl::detail::half_impl::half' diff --git a/clang/test/SemaSYCL/sampler.cpp b/clang/test/SemaSYCL/sampler.cpp index 461615fa3fa9a..032eec95c2c20 100644 --- a/clang/test/SemaSYCL/sampler.cpp +++ b/clang/test/SemaSYCL/sampler.cpp @@ -1,22 +1,26 @@ -// RUN: %clang_cc1 -S -fsycl -fsycl-is-device -triple spir64 -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -S -fsycl -fsycl-is-device -internal-isystem %S/Inputs -triple spir64 -ast-dump -sycl-std=2020 %s | FileCheck %s -#include "Inputs/sycl.hpp" +// This test checks if the compiler correctly initializes the SYCL Sampler object when passed as a kernel argument. -template -__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) { - kernelFunc(); -} +#include "sycl.hpp" + +sycl::queue myQueue; int main() { - cl::sycl::sampler Sampler; - kernel([=]() { - Sampler.use(); + + sycl::sampler Sampler; + + myQueue.submit([&](sycl::handler &h) { + h.single_task([=] { + Sampler.use(); + }); }); + return 0; } // Check declaration of the test kernel -// CHECK: FunctionDecl {{.*}}use_kernel_for_test{{.*}} 'void (sampler_t)' +// CHECK: FunctionDecl {{.*}}SamplerLambda{{.*}} 'void (sampler_t)' // // Check parameters of the test kernel // CHECK: ParmVarDecl {{.*}} used [[_arg_sampler:[0-9a-zA-Z_]+]] 'sampler_t' @@ -24,7 +28,7 @@ int main() { // Check that sampler field of the test kernel object is initialized using __init method // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void (__ocl_sampler_t)' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::sampler':'cl::sycl::sampler' lvalue +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::sampler':'sycl::sampler' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at {{.*}}sampler.cpp{{.*}})' lvalue Var {{.*}} '(lambda at {{.*}}sampler.cpp{{.*}})' // // Check the parameters of __init method diff --git a/clang/test/SemaSYCL/spec-const-kernel-arg.cpp b/clang/test/SemaSYCL/spec-const-kernel-arg.cpp index 486cc5d8ef525..a2bf4caa96c68 100644 --- a/clang/test/SemaSYCL/spec-const-kernel-arg.cpp +++ b/clang/test/SemaSYCL/spec-const-kernel-arg.cpp @@ -1,29 +1,33 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s // This test checks that compiler generates correct initialization for spec // constants -#include "Inputs/sycl.hpp" +#include "sycl.hpp" + +sycl::queue myQueue; struct SpecConstantsWrapper { - cl::sycl::ONEAPI::experimental::spec_constant SC1; - cl::sycl::ONEAPI::experimental::spec_constant SC2; + sycl::ONEAPI::experimental::spec_constant SC1; + sycl::ONEAPI::experimental::spec_constant SC2; }; int main() { - cl::sycl::ONEAPI::experimental::spec_constant SC; - SpecConstantsWrapper W; - cl::sycl::kernel_single_task( - [=]() { - (void)SC; - (void)W; - }); + sycl::ONEAPI::experimental::spec_constant SC; + SpecConstantsWrapper SCWrapper; + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + (void)SC; + (void)SCWrapper; + }); + }); } // CHECK: FunctionDecl {{.*}}kernel_sc{{.*}} 'void ()' // CHECK: VarDecl {{.*}}'(lambda at {{.*}}' // CHECK-NEXT: InitListExpr {{.*}}'(lambda at {{.*}}' -// CHECK-NEXT: CXXConstructExpr {{.*}}'cl::sycl::ONEAPI::experimental::spec_constant':'cl::sycl::ONEAPI::experimental::spec_constant' +// CHECK-NEXT: CXXConstructExpr {{.*}}'sycl::ONEAPI::experimental::spec_constant':'sycl::ONEAPI::experimental::spec_constant' // CHECK-NEXT: InitListExpr {{.*}} 'SpecConstantsWrapper' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::ONEAPI::experimental::spec_constant':'cl::sycl::ONEAPI::experimental::spec_constant' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::ONEAPI::experimental::spec_constant':'cl::sycl::ONEAPI::experimental::spec_constant' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::ONEAPI::experimental::spec_constant':'sycl::ONEAPI::experimental::spec_constant' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::ONEAPI::experimental::spec_constant':'sycl::ONEAPI::experimental::spec_constant' diff --git a/clang/test/SemaSYCL/streams.cpp b/clang/test/SemaSYCL/streams.cpp index 15c1c8625a7ad..bae0671bf1358 100644 --- a/clang/test/SemaSYCL/streams.cpp +++ b/clang/test/SemaSYCL/streams.cpp @@ -1,19 +1,18 @@ -// RUN: %clang_cc1 -S -fsycl -fsycl-is-device -triple spir64 -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -S -fsycl -fsycl-is-device -internal-isystem %S/Inputs -triple spir64 -ast-dump -sycl-std=2020 %s | FileCheck %s -#include "Inputs/sycl.hpp" +// This test demonstrates passing of SYCL stream instances as kernel arguments and checks if the compiler generates +// the correct ast-dump -template -__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) { - kernelFunc(); -} +#include "sycl.hpp" -using namespace cl::sycl; +sycl::queue myQueue; -handler H; +sycl::handler H; struct HasStreams { - stream s1{0, 0, H}; - stream s_array[2] = {{0, 0, H}, {0, 0, H}}; + sycl::stream s1{0, 0, H}; // stream(totalBufferSize, workItemBufferSize, handler) + + sycl::stream s_array[2] = {{0, 0, H}, {0, 0, H}}; }; struct HasArrayOfHasStreams { @@ -21,87 +20,92 @@ struct HasArrayOfHasStreams { HasStreams hs[2]; }; -void use() { - stream in_lambda{0, 0, H}; - stream in_lambda_array[2] = {{0, 0, H}, {0, 0, H}}; - stream in_lambda_mdarray[2][2] = {{{0, 0, H}, {0, 0, H}}, {{0, 0, H}, {0, 0, H}}}; +int main() { + sycl::stream in_lambda{0, 0, H}; + sycl::stream in_lambda_array[2] = {{0, 0, H}, {0, 0, H}}; + sycl::stream in_lambda_mdarray[2][2] = {{{0, 0, H}, {0, 0, H}}, {{0, 0, H}, {0, 0, H}}}; HasStreams Struct; HasArrayOfHasStreams haohs; HasArrayOfHasStreams haohs_array[2]; - kernel([=]() { - in_lambda.use(); - in_lambda_array[1].use(); - in_lambda_mdarray[1][1].use(); + myQueue.submit([&](sycl::handler &h) { + h.single_task([=]() { + in_lambda.use(); + in_lambda_array[1].use(); + in_lambda_mdarray[1][1].use(); - Struct.s1.use(); + Struct.s1.use(); - haohs.hs[0].s1.use(); - haohs_array[0].hs[0].s1.use(); + haohs.hs[0].s1.use(); + haohs_array[0].hs[0].s1.use(); + }); }); + + return 0; } // Function Declaration // CHECK: FunctionDecl {{.*}}stream_test{{.*}} // Initializers: + // CHECK: InitListExpr {{.*}} '(lambda at // 'in_lambda' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // 'in_lambda_array' -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // 'in_lambda_mdarray' -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2][2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2][2]' // sub-array 0 -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // sub-array 1 -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams struct // CHECK-NEXT: InitListExpr {{.*}} 'HasStreams' // HasStreams::s1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams::s_array -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasArrayOfHasStreams // CHECK-NEXT: InitListExpr {{.*}} 'HasArrayOfHasStreams' @@ -113,35 +117,35 @@ void use() { // HasStreams struct // CHECK-NEXT: InitListExpr {{.*}} 'HasStreams' // HasStreams::s1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams::s_array -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams struct // CHECK-NEXT: InitListExpr {{.*}} 'HasStreams' // HasStreams::s1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams::s_array -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasArrayOfHasStreams Array // CHECK-NEXT: InitListExpr {{.*}} 'HasArrayOfHasStreams [2]' @@ -154,35 +158,35 @@ void use() { // HasStreams struct // CHECK-NEXT: InitListExpr {{.*}} 'HasStreams' // HasStreams::s1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams::s_array -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams struct // CHECK-NEXT: InitListExpr {{.*}} 'HasStreams' // HasStreams::s1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams::s_array -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // CHECK-NEXT: InitListExpr {{.*}} 'HasArrayOfHasStreams' // HasArrayOfHasStreams::i // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' @@ -192,222 +196,222 @@ void use() { // HasStreams struct // CHECK-NEXT: InitListExpr {{.*}} 'HasStreams' // HasStreams::s1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams::s_array -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams struct // CHECK-NEXT: InitListExpr {{.*}} 'HasStreams' // HasStreams::s1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // HasStreams::s_array -// CHECK-NEXT: InitListExpr {{.*}} 'cl::sycl::stream [2]' +// CHECK-NEXT: InitListExpr {{.*}} 'sycl::stream [2]' // element 0 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // element 1 -// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::stream' 'void (const cl::sycl::stream &) noexcept' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::stream' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar +// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar // Calls to Init, note that the accessor in the stream comes first, since the // stream __init call depends on the accessor's call already having happened. // in_lambda __init // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue . +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue . +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // _in_lambda_array // element 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // _in_lambda_mdarray // [0][0] // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // [0][1] // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // [1][0] // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // [1][1] // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // HasStreams // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 @@ -415,9 +419,9 @@ void use() { // HasArrayOfHasStreams // First element // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -426,7 +430,7 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -435,11 +439,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -449,9 +453,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -461,11 +465,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -475,9 +479,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -487,9 +491,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // second element // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -498,7 +502,7 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -507,11 +511,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -521,9 +525,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -533,11 +537,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -547,9 +551,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -561,9 +565,9 @@ void use() { // HasArrayOfHasStreams array // First element // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -575,7 +579,7 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -587,11 +591,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -604,9 +608,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -619,11 +623,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -636,9 +640,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -651,9 +655,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // second element // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -665,7 +669,7 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -677,11 +681,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -694,9 +698,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -709,11 +713,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -726,9 +730,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -741,9 +745,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // second element // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -755,7 +759,7 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -767,11 +771,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -784,9 +788,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -799,11 +803,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -816,9 +820,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -831,9 +835,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // second element // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -845,7 +849,7 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -857,11 +861,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -874,9 +878,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -889,11 +893,11 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' -// CHECK-NEXT: MemberExpr {{.*}} 'void (cl::sycl::accessor{{.*}})' lvalue .__init +// CHECK-NEXT: MemberExpr {{.*}} 'void (sycl::accessor{{.*}})' lvalue .__init // CHECK-NEXT: MemberExpr {{.*}} 'accessor<{{.*}}' lvalue .acc -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -906,9 +910,9 @@ void use() { // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__init -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -924,24 +928,24 @@ void use() { // in_lambda __finalize // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue . +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // _in_lambda_array // element 0 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 @@ -949,44 +953,44 @@ void use() { // [0][0] // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // [0][1] // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // [1][0] // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // [1][1] // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 @@ -994,24 +998,24 @@ void use() { // HasStreams // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 0 // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1 @@ -1020,7 +1024,7 @@ void use() { // First element // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1030,9 +1034,9 @@ void use() { // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1043,9 +1047,9 @@ void use() { // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1056,7 +1060,7 @@ void use() { // second element // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1066,9 +1070,9 @@ void use() { // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1079,9 +1083,9 @@ void use() { // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1094,7 +1098,7 @@ void use() { // First element // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1107,9 +1111,9 @@ void use() { // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1123,9 +1127,9 @@ void use() { // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1139,7 +1143,7 @@ void use() { // second element // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1152,9 +1156,9 @@ void use() { // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1168,9 +1172,9 @@ void use() { // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1184,7 +1188,7 @@ void use() { // second element // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1197,9 +1201,9 @@ void use() { // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1213,9 +1217,9 @@ void use() { // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1229,7 +1233,7 @@ void use() { // second element // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream' lvalue .s1 +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue .s1 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1242,9 +1246,9 @@ void use() { // array: // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs @@ -1258,9 +1262,9 @@ void use() { // element 1 // CHECK: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ()' lvalue .__finalize -// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream' lvalue -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream *' -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array +// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs diff --git a/clang/test/SemaSYCL/wrapped-accessor.cpp b/clang/test/SemaSYCL/wrapped-accessor.cpp index f087621e3e575..e5825fac768f8 100644 --- a/clang/test/SemaSYCL/wrapped-accessor.cpp +++ b/clang/test/SemaSYCL/wrapped-accessor.cpp @@ -1,62 +1,64 @@ -// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s +// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s // This test checks that compiler generates correct kernel wrapper in case when // accessor is wrapped. -#include "Inputs/sycl.hpp" +#include "sycl.hpp" + +sycl::queue myQueue; template struct AccWrapper { Acc accessor; }; -template -__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) { - kernelFunc(); -} - int main() { - cl::sycl::accessor acc; + sycl::accessor acc; auto acc_wrapped = AccWrapper{acc}; - kernel( - [=]() { - acc_wrapped.accessor.use(); - }); + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + acc_wrapped.accessor.use(); + }); + }); + + return 0; } // Check declaration of the kernel -// CHECK: wrapped_access{{.*}} 'void (__global int *, cl::sycl::range<1>, cl::sycl::range<1>, cl::sycl::id<1>)' +// CHECK: wrapped_access{{.*}} 'void (__global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>)' // Check parameters of the kernel // CHECK: ParmVarDecl {{.*}} used _arg_accessor '__global int *' -// CHECK: ParmVarDecl {{.*}} used [[_arg_AccessRange:[0-9a-zA-Z_]+]] 'cl::sycl::range<1>' -// CHECK: ParmVarDecl {{.*}} used [[_arg_MemRange:[0-9a-zA-Z_]+]] 'cl::sycl::range<1>' -// CHECK: ParmVarDecl {{.*}} used [[_arg_Offset:[0-9a-zA-Z_]+]] 'cl::sycl::id<1>' +// CHECK: ParmVarDecl {{.*}} used [[_arg_AccessRange:[0-9a-zA-Z_]+]] 'sycl::range<1>' +// CHECK: ParmVarDecl {{.*}} used [[_arg_MemRange:[0-9a-zA-Z_]+]] 'sycl::range<1>' +// CHECK: ParmVarDecl {{.*}} used [[_arg_Offset:[0-9a-zA-Z_]+]] 'sycl::id<1>' // Check that wrapper object itself is initialized with corresponding kernel // argument // CHECK: VarDecl {{.*}}'(lambda at {{.*}}wrapped-accessor.cpp{{.*}})' // CHECK-NEXT: InitListExpr {{.*}}'(lambda at {{.*}}wrapped-accessor.cpp{{.*}})' -// CHECK-NEXT: InitListExpr {{.*}}'AccWrapper>' -// CHECK-NEXT: CXXConstructExpr {{.*}}'cl::sycl::accessor':'cl::sycl::accessor' 'void () noexcept' +// CHECK-NEXT: InitListExpr {{.*}}'AccWrapper>' +// CHECK-NEXT: CXXConstructExpr {{.*}}'sycl::accessor':'sycl::accessor' 'void () noexcept' // Check that accessor field of the wrapper object is initialized using __init method // CHECK-NEXT: CXXMemberCallExpr {{.*}} 'void' // CHECK-NEXT: MemberExpr {{.*}} 'void ({{.*}}PtrType, range<1>, range<1>, id<1>)' lvalue .__init -// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::accessor':'cl::sycl::accessor' lvalue .accessor {{.*}} -// CHECK-NEXT: MemberExpr {{.*}} 'AccWrapper>':'AccWrapper>' lvalue . +// CHECK-NEXT: MemberExpr {{.*}} 'sycl::accessor':'sycl::accessor' lvalue .accessor {{.*}} +// CHECK-NEXT: MemberExpr {{.*}} 'AccWrapper>':'AccWrapper>' lvalue . // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at {{.*}}wrapped-accessor.cpp{{.*}})' lvalue Var {{.*}} '(lambda at {{.*}}wrapped-accessor.cpp{{.*}})' // Parameters of the _init method // CHECK-NEXT: ImplicitCastExpr {{.*}} // CHECK-NEXT: DeclRefExpr {{.*}} '__global int *' lvalue ParmVar {{.*}} '_arg_accessor' '__global int *' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'range<1>':'cl::sycl::range<1>' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::range<1>' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::range<1>' lvalue ParmVar {{.*}} '[[_arg_AccessRange]]' 'cl::sycl::range<1>' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'range<1>':'sycl::range<1>' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::range<1>' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::range<1>' lvalue ParmVar {{.*}} '[[_arg_AccessRange]]' 'sycl::range<1>' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'range<1>':'cl::sycl::range<1>' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::range<1>' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::range<1>' lvalue ParmVar {{.*}} '[[_arg_MemRange]]' 'cl::sycl::range<1>' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'range<1>':'sycl::range<1>' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::range<1>' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::range<1>' lvalue ParmVar {{.*}} '[[_arg_MemRange]]' 'sycl::range<1>' -// CHECK-NEXT: CXXConstructExpr {{.*}} 'id<1>':'cl::sycl::id<1>' -// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const cl::sycl::id<1>' lvalue -// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::id<1>' lvalue ParmVar {{.*}} '[[_arg_Offset]]' 'cl::sycl::id<1>' +// CHECK-NEXT: CXXConstructExpr {{.*}} 'id<1>':'sycl::id<1>' +// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::id<1>' lvalue +// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::id<1>' lvalue ParmVar {{.*}} '[[_arg_Offset]]' 'sycl::id<1>'