From 3207cc224649ad57b56c454f39b94b2e5c28e03e Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Tue, 5 Jan 2021 19:14:20 -0800
Subject: [PATCH 01/22] [SYCL] Update FE tests to have a common infrastructure

---
 .../test/SemaSYCL/accessors-targets-image.cpp | 77 +++++++++++--------
 1 file changed, 47 insertions(+), 30 deletions(-)

diff --git a/clang/test/SemaSYCL/accessors-targets-image.cpp b/clang/test/SemaSYCL/accessors-targets-image.cpp
index 5afa2010451cb..f687c7527a55c 100644
--- a/clang/test/SemaSYCL/accessors-targets-image.cpp
+++ b/clang/test/SemaSYCL/accessors-targets-image.cpp
@@ -1,66 +1,83 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s
+// RUN: %clang_cc1 -fsycl -fsycl-is-device -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"
 
 using namespace cl::sycl;
 
-template <typename name, typename Func>
-__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) {
-  kernelFunc();
-}
+queue q;
 
 int main() {
 
+  // 1-dimensional accessor with Read-only access
   accessor<int, 1, access::mode::read,
            access::target::image, access::placeholder::false_t>
       image_acc1d_read;
-  kernel<class use_image1d_r>(
-      [=]() {
-        image_acc1d_read.use();
-      });
 
+  q.submit([&](handler &h) {
+    h.single_task<class use_image1d_r>(
+        [=]() {
+          image_acc1d_read.use();
+        });
+  });
+
+  // 2-dimensional accessor with Read-only access
   accessor<int, 2, access::mode::read,
            access::target::image, access::placeholder::false_t>
       image_acc2d_read;
-  kernel<class use_image2d_r>(
-      [=]() {
-        image_acc2d_read.use();
-      });
+  q.submit([&](handler &h) {
+    h.single_task<class use_image2d_r>(
+        [=]() {
+          image_acc2d_read.use();
+        });
+  });
 
+  // 3-dimensional accessor with Read-only access
   accessor<int, 3, access::mode::read,
            access::target::image, access::placeholder::false_t>
       image_acc3d_read;
-  kernel<class use_image3d_r>(
-      [=]() {
-        image_acc3d_read.use();
-      });
 
+  q.submit([&](handler &h) {
+    h.single_task<class use_image3d_r>(
+        [=]() {
+          image_acc3d_read.use();
+        });
+  });
+
+  // 1-dimensional accessor with Write-only access
   accessor<int, 1, access::mode::write,
            access::target::image, access::placeholder::false_t>
       image_acc1d_write;
-  kernel<class use_image1d_w>(
-      [=]() {
-        image_acc1d_write.use();
-      });
+  q.submit([&](handler &h) {
+    h.single_task<class use_image1d_w>(
+        [=]() {
+          image_acc1d_write.use();
+        });
+  });
 
+  // 2-dimensional accessor with Write-only access
   accessor<int, 2, access::mode::write,
            access::target::image, access::placeholder::false_t>
       image_acc2d_write;
-  kernel<class use_image2d_w>(
-      [=]() {
-        image_acc2d_write.use();
-      });
+  q.submit([&](handler &h) {
+    h.single_task<class use_image2d_w>(
+        [=]() {
+          image_acc2d_write.use();
+        });
+  });
 
+  // 3-dimensional accessor with Write-only access
   accessor<int, 3, access::mode::write,
            access::target::image, access::placeholder::false_t>
       image_acc3d_write;
-  kernel<class use_image3d_w>(
-      [=]() {
-        image_acc3d_write.use();
-      });
+  q.submit([&](handler &h) {
+    h.single_task<class use_image3d_w>(
+        [=]() {
+          image_acc3d_write.use();
+        });
+  });
 }
 
 // CHECK: {{.*}}use_image1d_r 'void (__read_only image1d_t)'

From 39966fcb236fb24f8e617dbc1f26a2ec6478805b Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Wed, 6 Jan 2021 10:26:51 -0800
Subject: [PATCH 02/22] Avoid using "cl::sycl" namespace

---
 .../test/SemaSYCL/accessors-targets-image.cpp | 52 +++++++++----------
 1 file changed, 25 insertions(+), 27 deletions(-)

diff --git a/clang/test/SemaSYCL/accessors-targets-image.cpp b/clang/test/SemaSYCL/accessors-targets-image.cpp
index f687c7527a55c..335d0c0a5adc1 100644
--- a/clang/test/SemaSYCL/accessors-targets-image.cpp
+++ b/clang/test/SemaSYCL/accessors-targets-image.cpp
@@ -5,76 +5,74 @@
 
 #include "Inputs/sycl.hpp"
 
-using namespace cl::sycl;
-
-queue q;
+cl::sycl::queue q;
 
 int main() {
 
   // 1-dimensional accessor with Read-only access
-  accessor<int, 1, access::mode::read,
-           access::target::image, access::placeholder::false_t>
+  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read,
+                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
       image_acc1d_read;
 
-  q.submit([&](handler &h) {
+  q.submit([&](cl::sycl::handler &h) {
     h.single_task<class use_image1d_r>(
-        [=]() {
+        [=] {
           image_acc1d_read.use();
         });
   });
 
   // 2-dimensional accessor with Read-only access
-  accessor<int, 2, access::mode::read,
-           access::target::image, access::placeholder::false_t>
+  cl::sycl::accessor<int, 2, cl::sycl::access::mode::read,
+                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
       image_acc2d_read;
-  q.submit([&](handler &h) {
+  q.submit([&](cl::sycl::handler &h) {
     h.single_task<class use_image2d_r>(
-        [=]() {
+        [=] {
           image_acc2d_read.use();
         });
   });
 
   // 3-dimensional accessor with Read-only access
-  accessor<int, 3, access::mode::read,
-           access::target::image, access::placeholder::false_t>
+  cl::sycl::accessor<int, 3, cl::sycl::access::mode::read,
+                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
       image_acc3d_read;
 
-  q.submit([&](handler &h) {
+  q.submit([&](cl::sycl::handler &h) {
     h.single_task<class use_image3d_r>(
-        [=]() {
+        [=] {
           image_acc3d_read.use();
         });
   });
 
   // 1-dimensional accessor with Write-only access
-  accessor<int, 1, access::mode::write,
-           access::target::image, access::placeholder::false_t>
+  cl::sycl::accessor<int, 1, cl::sycl::access::mode::write,
+                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
       image_acc1d_write;
-  q.submit([&](handler &h) {
+  q.submit([&](cl::sycl::handler &h) {
     h.single_task<class use_image1d_w>(
-        [=]() {
+        [=] {
           image_acc1d_write.use();
         });
   });
 
   // 2-dimensional accessor with Write-only access
-  accessor<int, 2, access::mode::write,
-           access::target::image, access::placeholder::false_t>
+  cl::sycl::accessor<int, 2, cl::sycl::access::mode::write,
+                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
       image_acc2d_write;
-  q.submit([&](handler &h) {
+  q.submit([&](cl::sycl::handler &h) {
     h.single_task<class use_image2d_w>(
-        [=]() {
+        [=] {
           image_acc2d_write.use();
         });
   });
 
   // 3-dimensional accessor with Write-only access
-  accessor<int, 3, access::mode::write,
-           access::target::image, access::placeholder::false_t>
+  cl::sycl::accessor<int, 3, cl::sycl::access::mode::write,
+                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
       image_acc3d_write;
-  q.submit([&](handler &h) {
+  q.submit([&](cl::sycl::handler &h) {
     h.single_task<class use_image3d_w>(
-        [=]() {
+        [=] {
           image_acc3d_write.use();
         });
   });

From 2d69057547ca863762fa7bb3ecb5ad32c18d9888 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Wed, 6 Jan 2021 10:58:52 -0800
Subject: [PATCH 03/22] Update accessors-targets.cpp

---
 clang/test/SemaSYCL/accessors-targets.cpp | 55 ++++++++++++++---------
 1 file changed, 33 insertions(+), 22 deletions(-)

diff --git a/clang/test/SemaSYCL/accessors-targets.cpp b/clang/test/SemaSYCL/accessors-targets.cpp
index 60b5b9159fc14..0d6026be21dd1 100644
--- a/clang/test/SemaSYCL/accessors-targets.cpp
+++ b/clang/test/SemaSYCL/accessors-targets.cpp
@@ -1,11 +1,11 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s
+// RUN: %clang_cc1 -fsycl -fsycl-is-device -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"
 
-using namespace cl::sycl;
+cl::sycl::queue q;
 
 template <typename name, typename Func>
 __attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) {
@@ -13,28 +13,39 @@ __attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) {
 }
 
 int main() {
-
-  accessor<int, 1, access::mode::read_write,
-           access::target::local>
+  // Access work-group local memory with read and write access.
+  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write,
+                     cl::sycl::access::target::local>
       local_acc;
-  accessor<int, 1, access::mode::read_write,
-           access::target::global_buffer>
+  // Access buffer via global memory with read and write access.
+  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write,
+                     cl::sycl::access::target::global_buffer>
       global_acc;
-  accessor<int, 1, access::mode::read_write,
-           access::target::constant_buffer>
+  // Access buffer via constant memory with read and write access.
+  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write,
+                     cl::sycl::access::target::constant_buffer>
       constant_acc;
-  kernel<class use_local>(
-      [=]() {
-        local_acc.use();
-      });
-  kernel<class use_global>(
-      [=]() {
-        global_acc.use();
-      });
-  kernel<class use_constant>(
-      [=]() {
-        constant_acc.use();
-      });
+
+  q.submit([&](cl::sycl::handler &h) {
+    h.single_task<class use_local>(
+        [=] {
+          local_acc.use();
+        });
+  });
+
+  q.submit([&](cl::sycl::handler &h) {
+    h.single_task<class use_global>(
+        [=] {
+          global_acc.use();
+        });
+  });
+
+  q.submit([&](cl::sycl::handler &h) {
+    h.single_task<class use_constant>(
+        [=] {
+          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>)'

From bbbba5f9db347339d10d42849871944a814740e9 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Wed, 6 Jan 2021 11:10:25 -0800
Subject: [PATCH 04/22] Update allow-constexpr-recursion.cpp

---
 .../test/SemaSYCL/allow-constexpr-recursion.cpp  | 16 ++++++++++------
 1 file changed, 10 insertions(+), 6 deletions(-)

diff --git a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
index 6b17af6351e22..874c418b0b5f0 100644
--- a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
+++ b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
@@ -1,9 +1,8 @@
 // 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
 
-template <typename name, typename Func>
-__attribute__((sycl_kernel)) void kernel_single_task(const Func &kernelFunc) {
-  kernelFunc();
-}
+#include "Inputs/sycl.hpp"
+
+cl::sycl::queue q;
 
 // expected-note@+1{{function implemented using recursion declared here}}
 constexpr int constexpr_recurse1(int n);
@@ -71,6 +70,11 @@ void constexpr_recurse_test_err() {
 }
 
 int main() {
-  kernel_single_task<class fake_kernel>([]() { constexpr_recurse_test(); });
-  kernel_single_task<class fake_kernel>([]() { constexpr_recurse_test_err(); });
+  q.submit([&](cl::sycl::handler &h) {
+    h.single_task<class fake_kernel>([]() { constexpr_recurse_test(); });
+  });
+
+  q.submit([&](cl::sycl::handler &h) {
+    h.single_task<class fake_kernel>([]() { constexpr_recurse_test_err(); });
+  });
 }

From 04b1200c561487b7eecb0f2e5981eb049db58326 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Wed, 6 Jan 2021 11:16:57 -0800
Subject: [PATCH 05/22] Remove unused code

---
 clang/test/SemaSYCL/accessors-targets.cpp | 5 -----
 1 file changed, 5 deletions(-)

diff --git a/clang/test/SemaSYCL/accessors-targets.cpp b/clang/test/SemaSYCL/accessors-targets.cpp
index 0d6026be21dd1..043e0d1f5e8c1 100644
--- a/clang/test/SemaSYCL/accessors-targets.cpp
+++ b/clang/test/SemaSYCL/accessors-targets.cpp
@@ -7,11 +7,6 @@
 
 cl::sycl::queue q;
 
-template <typename name, typename Func>
-__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) {
-  kernelFunc();
-}
-
 int main() {
   // Access work-group local memory with read and write access.
   cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write,

From 2a169a4f864d81fe89c4aac86563f9a08eac9528 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Wed, 6 Jan 2021 14:18:48 -0800
Subject: [PATCH 06/22] Update array-kernel-param-neg.cpp

---
 .../SemaSYCL/allow-constexpr-recursion.cpp    |  2 +
 .../test/SemaSYCL/array-kernel-param-neg.cpp  | 52 +++++++++++--------
 2 files changed, 31 insertions(+), 23 deletions(-)

diff --git a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
index 874c418b0b5f0..dc4e1a9dd0ecd 100644
--- a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
+++ b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
@@ -1,5 +1,7 @@
 // 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
 
+// This test verifies that a SYCL kernel executed on a device, cannot call a recursive function.
+
 #include "Inputs/sycl.hpp"
 
 cl::sycl::queue q;
diff --git a/clang/test/SemaSYCL/array-kernel-param-neg.cpp b/clang/test/SemaSYCL/array-kernel-param-neg.cpp
index 5ad8569feebe8..beaa41b1bbbc2 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 -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 "Inputs/sycl.hpp"
+
+cl::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 <typename Name, typename Func>
-__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<class kernel_capture_refs>([=] {
-    // 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([&](cl::sycl::handler &h) {
+    h.single_task<class kernel_capture_refs>([=] {
+      // 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<class kernel_bad_array>(es);
+  q.submit([&](cl::sycl::handler &h) {
+    // expected-error@+1 {{variable 'UnknownSizeArrayObj' with flexible array member cannot be captured in a lambda expression}}
+    h.single_task<class kernel_bad_array>(UnknownSizeArrayObj);
+  });
 }

From d1e96fa15d9991a1964681c2c11033807d7cc44e Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Thu, 7 Jan 2021 19:22:50 -0800
Subject: [PATCH 07/22] Inline "cl" namespace and update tests

---
 clang/test/SemaSYCL/Inputs/sycl.hpp           |  2 +-
 .../test/SemaSYCL/accessors-targets-image.cpp | 38 +++++++++----------
 clang/test/SemaSYCL/accessors-targets.cpp     | 26 ++++++-------
 .../SemaSYCL/allow-constexpr-recursion.cpp    |  6 +--
 .../test/SemaSYCL/array-kernel-param-neg.cpp  |  6 +--
 5 files changed, 39 insertions(+), 39 deletions(-)

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/accessors-targets-image.cpp b/clang/test/SemaSYCL/accessors-targets-image.cpp
index 335d0c0a5adc1..53888ad55930d 100644
--- a/clang/test/SemaSYCL/accessors-targets-image.cpp
+++ b/clang/test/SemaSYCL/accessors-targets-image.cpp
@@ -5,16 +5,16 @@
 
 #include "Inputs/sycl.hpp"
 
-cl::sycl::queue q;
+sycl::queue q;
 
 int main() {
 
   // 1-dimensional accessor with Read-only access
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read,
-                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
+  sycl::accessor<int, 1, sycl::access::mode::read,
+                 sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc1d_read;
 
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class use_image1d_r>(
         [=] {
           image_acc1d_read.use();
@@ -22,10 +22,10 @@ int main() {
   });
 
   // 2-dimensional accessor with Read-only access
-  cl::sycl::accessor<int, 2, cl::sycl::access::mode::read,
-                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
+  sycl::accessor<int, 2, sycl::access::mode::read,
+                 sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc2d_read;
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class use_image2d_r>(
         [=] {
           image_acc2d_read.use();
@@ -33,11 +33,11 @@ int main() {
   });
 
   // 3-dimensional accessor with Read-only access
-  cl::sycl::accessor<int, 3, cl::sycl::access::mode::read,
-                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
+  sycl::accessor<int, 3, sycl::access::mode::read,
+                 sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc3d_read;
 
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class use_image3d_r>(
         [=] {
           image_acc3d_read.use();
@@ -45,10 +45,10 @@ int main() {
   });
 
   // 1-dimensional accessor with Write-only access
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::write,
-                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
+  sycl::accessor<int, 1, sycl::access::mode::write,
+                 sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc1d_write;
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class use_image1d_w>(
         [=] {
           image_acc1d_write.use();
@@ -56,10 +56,10 @@ int main() {
   });
 
   // 2-dimensional accessor with Write-only access
-  cl::sycl::accessor<int, 2, cl::sycl::access::mode::write,
-                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
+  sycl::accessor<int, 2, sycl::access::mode::write,
+                 sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc2d_write;
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class use_image2d_w>(
         [=] {
           image_acc2d_write.use();
@@ -67,10 +67,10 @@ int main() {
   });
 
   // 3-dimensional accessor with Write-only access
-  cl::sycl::accessor<int, 3, cl::sycl::access::mode::write,
-                     cl::sycl::access::target::image, cl::sycl::access::placeholder::false_t>
+  sycl::accessor<int, 3, sycl::access::mode::write,
+                 sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc3d_write;
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class use_image3d_w>(
         [=] {
           image_acc3d_write.use();
diff --git a/clang/test/SemaSYCL/accessors-targets.cpp b/clang/test/SemaSYCL/accessors-targets.cpp
index 043e0d1f5e8c1..5d90adeb44700 100644
--- a/clang/test/SemaSYCL/accessors-targets.cpp
+++ b/clang/test/SemaSYCL/accessors-targets.cpp
@@ -5,43 +5,43 @@
 
 #include "Inputs/sycl.hpp"
 
-cl::sycl::queue q;
+sycl::queue q;
 
 int main() {
   // Access work-group local memory with read and write access.
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write,
-                     cl::sycl::access::target::local>
+  sycl::accessor<int, 1, sycl::access::mode::read_write,
+                 sycl::access::target::local>
       local_acc;
   // Access buffer via global memory with read and write access.
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write,
-                     cl::sycl::access::target::global_buffer>
+  sycl::accessor<int, 1, sycl::access::mode::read_write,
+                 sycl::access::target::global_buffer>
       global_acc;
   // Access buffer via constant memory with read and write access.
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write,
-                     cl::sycl::access::target::constant_buffer>
+  sycl::accessor<int, 1, sycl::access::mode::read_write,
+                 sycl::access::target::constant_buffer>
       constant_acc;
 
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class use_local>(
         [=] {
           local_acc.use();
         });
   });
 
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class use_global>(
         [=] {
           global_acc.use();
         });
   });
 
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class use_constant>(
         [=] {
           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 dc4e1a9dd0ecd..5c5542a61c077 100644
--- a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
+++ b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
@@ -4,7 +4,7 @@
 
 #include "Inputs/sycl.hpp"
 
-cl::sycl::queue q;
+sycl::queue q;
 
 // expected-note@+1{{function implemented using recursion declared here}}
 constexpr int constexpr_recurse1(int n);
@@ -72,11 +72,11 @@ void constexpr_recurse_test_err() {
 }
 
 int main() {
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class fake_kernel>([]() { constexpr_recurse_test(); });
   });
 
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class fake_kernel>([]() { 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 beaa41b1bbbc2..c16c22081e16e 100755
--- a/clang/test/SemaSYCL/array-kernel-param-neg.cpp
+++ b/clang/test/SemaSYCL/array-kernel-param-neg.cpp
@@ -6,7 +6,7 @@
 
 #include "Inputs/sycl.hpp"
 
-cl::sycl::queue q;
+sycl::queue q;
 
 struct NonTrivialCopyStruct {
   int i;
@@ -33,7 +33,7 @@ void test() {
   // expected-note@+1 {{'UnknownSizeArrayObj' declared here}}
   Array UnknownSizeArrayObj;
 
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     h.single_task<class kernel_capture_refs>([=] {
       // expected-error@+1 {{kernel parameter has non-trivially copy constructible class/struct type}}
       int b = NTCSObject[2].i;
@@ -42,7 +42,7 @@ void test() {
     });
   });
 
-  q.submit([&](cl::sycl::handler &h) {
+  q.submit([&](sycl::handler &h) {
     // expected-error@+1 {{variable 'UnknownSizeArrayObj' with flexible array member cannot be captured in a lambda expression}}
     h.single_task<class kernel_bad_array>(UnknownSizeArrayObj);
   });

From 4f006e751f7e80979450aa5ae4f9af62a8d195d4 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Fri, 8 Jan 2021 11:26:53 -0800
Subject: [PATCH 08/22] Update basic-kernel-wrapper.cpp

---
 clang/test/SemaSYCL/basic-kernel-wrapper.cpp | 51 +++++++++-----------
 1 file changed, 24 insertions(+), 27 deletions(-)

diff --git a/clang/test/SemaSYCL/basic-kernel-wrapper.cpp b/clang/test/SemaSYCL/basic-kernel-wrapper.cpp
index f1ff38ec993df..6bdbf3dc6fe78 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 -ast-dump -sycl-std=2020 %s | FileCheck %s
 
 // This test checks that compiler generates correct kernel wrapper for basic
 // case.
 
 #include "Inputs/sycl.hpp"
 
-template <typename Acc>
-struct AccWrapper { Acc accessor; };
-
-template <typename name, typename Func>
-__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) {
-  kernelFunc();
-}
+sycl::queue myQueue;
 
 int main() {
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write> acc;
-  kernel<class kernel_wrapper>(
-      [=]() {
-        acc.use();
-      });
+  sycl::accessor<int, 1, sycl::access::mode::read_write> readWriteAccessor;
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class kernel_wrapper>(
+        [=]() {
+          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<int, 1, cl::sycl::access::mode::read_write>':'cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>' lvalue .
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::accessor<int, 1, sycl::access::mode::read_write>':'sycl::accessor<int, 1, sycl::access::mode::read_write, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>' lvalue .
 // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at {{.*}}basic-kernel-wrapper.cpp{{.*}})' lvalue Var
 
 // CHECK-NEXT: ImplicitCastExpr {{.*}} <LValueToRValue>
 // 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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
 

From bbd4ff7d02703cdab3db69303f3486a62e496cc7 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Fri, 8 Jan 2021 11:37:36 -0800
Subject: [PATCH 09/22] Update half-kernel-arg.cpp

---
 clang/test/SemaSYCL/half-kernel-arg.cpp | 28 ++++++++++++++-----------
 1 file changed, 16 insertions(+), 12 deletions(-)

diff --git a/clang/test/SemaSYCL/half-kernel-arg.cpp b/clang/test/SemaSYCL/half-kernel-arg.cpp
index e6297351fad8b..b579bde840cca 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 -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"
 
+sycl::queue myQueue;
+
 int main() {
-  cl::sycl::half HostHalf;
-  cl::sycl::kernel_single_task<class kernel_half>(
-      [=]() {
-        cl::sycl::half KernelHalf = HostHalf;
-      });
+  sycl::half HostHalf;
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class kernel_half>(
+        [=]() {
+          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'

From ecccf60d2f780692297e509f6128a4304f55bfdb Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Mon, 11 Jan 2021 10:38:04 -0800
Subject: [PATCH 10/22] Update fake-accessors.cpp

---
 clang/test/SemaSYCL/fake-accessors.cpp | 62 +++++++++++++-------------
 1 file changed, 31 insertions(+), 31 deletions(-)

diff --git a/clang/test/SemaSYCL/fake-accessors.cpp b/clang/test/SemaSYCL/fake-accessors.cpp
index d4dd2a0f60209..fd1cc26d0e24e 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 -ast-dump -sycl-std=2020 %s | FileCheck %s
 
 #include "Inputs/sycl.hpp"
 
-namespace foo {
+sycl::queue myQueue;
+
+namespace fake {
 namespace cl {
 namespace sycl {
 class accessor {
@@ -11,46 +13,44 @@ class accessor {
 };
 } // namespace sycl
 } // namespace cl
-} // namespace foo
+} // namespace fake
 
 class accessor {
 public:
   int field;
 };
 
-typedef cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write,
-                           cl::sycl::access::target::global_buffer>
-    MyAccessorTD;
-
-using MyAccessorA = cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write,
-                                       cl::sycl::access::target::global_buffer>;
-
-template <typename name, typename Func>
-__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) {
-  kernelFunc();
-}
-
 int main() {
-  foo::cl::sycl::accessor acc = {1};
+  fake::cl::sycl::accessor FakeAccessor = {1};
   accessor acc1 = {1};
 
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write> accessorA;
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write> accessorB;
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write> accessorC;
-    kernel<class fake_accessors>(
-        [=]() {
-          accessorA.use((void*)(acc.field + acc1.field));
+  sycl::accessor<int, 1, sycl::access::mode::read_write> accessorA;
+  sycl::accessor<int, 1, sycl::access::mode::read_write> accessorB;
+  sycl::accessor<int, 1, sycl::access::mode::read_write> accessorC;
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class fake_accessors>(
+        [=] {
+          accessorA.use((void *)(FakeAccessor.field + acc1.field));
         });
-    kernel<class accessor_typedef>(
-        [=]() {
-          accessorB.use((void*)(acc.field + acc1.field));
+  });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class accessor_typedef>(
+        [=] {
+          accessorB.use((void *)(FakeAccessor.field + acc1.field));
         });
-    kernel<class accessor_alias>(
-        [=]() {
-          accessorC.use((void*)(acc.field + acc1.field));
+  });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class accessor_alias>(
+        [=] {
+          accessorC.use((void *)(FakeAccessor.field + acc1.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)

From 014f3e7113d3e345771778536eb20e5fe30f604b Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Mon, 11 Jan 2021 14:10:07 -0800
Subject: [PATCH 11/22] Update decomposition.cpp

---
 clang/test/SemaSYCL/decomposition.cpp | 168 +++++++++++++++-----------
 1 file changed, 100 insertions(+), 68 deletions(-)

diff --git a/clang/test/SemaSYCL/decomposition.cpp b/clang/test/SemaSYCL/decomposition.cpp
index 6830c944f239c..f871445057b00 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 -ast-dump -sycl-std=2020 %s | FileCheck %s
+
+// 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.
 
 #include "Inputs/sycl.hpp"
 
-using namespace cl::sycl;
+sycl::queue myQueue;
 
-struct has_acc {
-  accessor<char, 1, access::mode::read> acc;
+struct StructWithAccessor {
+  sycl::accessor<char, 1, sycl::access::mode::read> acc;
 };
 
-struct acc_base : accessor<char, 1, access::mode::read> {
+struct StructInheritedAccessor : sycl::accessor<char, 1, sycl::access::mode::read> {
   int i;
 };
 
-struct has_sampler {
-  sampler sampl;
+struct StructWithSampler {
+  sycl::sampler sampl;
 };
 
-struct has_spec_const {
-  ONEAPI::experimental::spec_constant<int, class f1> SC;
+struct StructWithSpecConst {
+  sycl::ONEAPI::experimental::spec_constant<int, class f1> 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 <typename T>
-struct Test1 {
+struct StructWithArray {
   T a;
   T b[2];
-  non_decomposed d;
+  StructNonDecomposed SimpleStruct;
   int i;
 };
 
 template <typename T>
-struct Test2 : T {
-  non_decomposed d;
+struct DerivedStruct : T {
+  StructNonDecomposed SimpleStruct;
   int i;
 };
 
-template <typename Name, typename Func>
-__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 d2;
+  StructWithNonDecomposedStruct d2s[5];
   // Check to ensure that these are not decomposed.
-  kernel<class NonDecomp>([=]() { 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<class NonDecomposed>([=]() { return SimpleStruct.i + ArrayOfSimpleStruct[0].i + d2.i + d2s[0].i; });
+  });
+  // CHECK: FunctionDecl {{.*}}NonDecomposed{{.*}} 'void (StructNonDecomposed, __wrapper_class, StructWithNonDecomposedStruct, __wrapper_class)'
 
   {
-    Test1<has_acc> t1;
-    kernel<class Acc1>([=]() { 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<has_acc> t2;
-    kernel<class Acc2>([=]() { 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<acc_base> t3;
-    kernel<class Acc3>([=]() { 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<acc_base> t4;
-    kernel<class Acc4>([=]() { 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<StructWithAccessor> t1;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Acc1>([=]() { 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<StructWithAccessor> t2;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Acc2>([=]() { return t2.i; });
+    });
+    // CHECK: FunctionDecl {{.*}}Acc2{{.*}} 'void (__global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, StructNonDecomposed, int)'
+
+    StructWithArray<StructInheritedAccessor> t3;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Acc3>([=]() { 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<StructInheritedAccessor> t4;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Acc4>([=]() { return t4.i; });
+    });
+    // CHECK: FunctionDecl {{.*}}Acc4{{.*}} 'void (__global char *, sycl::range<1>, sycl::range<1>, sycl::id<1>, int, StructNonDecomposed, int)'
   }
 
   {
-    Test1<has_sampler> t1;
-    kernel<class Sampl1>([=]() { return t1.i; });
-    // CHECK: FunctionDecl {{.*}}Sampl1{{.*}} 'void (sampler_t, sampler_t, sampler_t, non_decomposed, int)'
-    Test2<has_sampler> t2;
-    kernel<class Sampl2>([=]() { return t2.i; });
-    // CHECK: FunctionDecl {{.*}}Sampl2{{.*}} 'void (sampler_t, non_decomposed, int)'
+    StructWithArray<StructWithSampler> t1;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Sampl1>([=]() { return t1.i; });
+    });
+    // CHECK: FunctionDecl {{.*}}Sampl1{{.*}} 'void (sampler_t, sampler_t, sampler_t, StructNonDecomposed, int)'
+
+    DerivedStruct<StructWithSampler> t2;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Sampl2>([=]() { return t2.i; });
+    });
+    // CHECK: FunctionDecl {{.*}}Sampl2{{.*}} 'void (sampler_t, StructNonDecomposed, int)'
   }
 
   {
-    Test1<has_spec_const> t1;
-    kernel<class SpecConst1>([=]() { return t1.i; });
-    // CHECK: FunctionDecl {{.*}}SpecConst{{.*}} 'void (non_decomposed, int)'
-    Test2<has_spec_const> t2;
-    kernel<class SpecConst2>([=]() { return t2.i; });
-    // CHECK: FunctionDecl {{.*}}SpecConst2{{.*}} 'void (non_decomposed, int)'
+    StructWithArray<StructWithSpecConst> t1;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class SpecConst1>([=]() { return t1.i; });
+    });
+    // CHECK: FunctionDecl {{.*}}SpecConst{{.*}} 'void (StructNonDecomposed, int)'
+
+    DerivedStruct<StructWithSpecConst> t2;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class SpecConst2>([=]() { return t2.i; });
+    });
+    // CHECK: FunctionDecl {{.*}}SpecConst2{{.*}} 'void (StructNonDecomposed, int)'
   }
 
   {
-    Test1<has_stream> t1;
-    kernel<class Stream1>([=]() { 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<has_stream> t2;
-    kernel<class Stream2>([=]() { 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<StructWithStream> t1;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Stream1>([=]() { 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<StructWithStream> t2;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Stream2>([=]() { return t2.i; });
+    });
+    // CHECK: FunctionDecl {{.*}}Stream2{{.*}} 'void (sycl::stream, __global int *, sycl::range<1>, sycl::range<1>, sycl::id<1>, StructNonDecomposed, int)'
   }
 
   {
-    Test1<has_half> t1;
-    kernel<class Half1>([=]() { return t1.i; });
-    // CHECK: FunctionDecl {{.*}}Half1{{.*}} 'void (cl::sycl::half, cl::sycl::half, cl::sycl::half, non_decomposed, int)'
-    Test2<has_half> t2;
-    kernel<class Half2>([=]() { return t2.i; });
-    // CHECK: FunctionDecl {{.*}}Half2{{.*}} 'void (cl::sycl::half, non_decomposed, int)'
+    StructWithArray<StructWithHalf> t1;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Half1>([=]() { return t1.i; });
+    });
+    // CHECK: FunctionDecl {{.*}}Half1{{.*}} 'void (sycl::half, sycl::half, sycl::half, StructNonDecomposed, int)'
+
+    DerivedStruct<StructWithHalf> t2;
+    myQueue.submit([&](sycl::handler &h) {
+      h.single_task<class Half2>([=]() { return t2.i; });
+    });
+    // CHECK: FunctionDecl {{.*}}Half2{{.*}} 'void (sycl::half, StructNonDecomposed, int)'
   }
 }

From 0a916bff1e9d48d84254691f31c1475515064e20 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Mon, 11 Jan 2021 18:21:22 -0800
Subject: [PATCH 12/22] Update spec-const-kernel-arg.cpp

---
 clang/test/SemaSYCL/spec-const-kernel-arg.cpp | 20 +++++++++----------
 1 file changed, 10 insertions(+), 10 deletions(-)

diff --git a/clang/test/SemaSYCL/spec-const-kernel-arg.cpp b/clang/test/SemaSYCL/spec-const-kernel-arg.cpp
index 486cc5d8ef525..240241062ccb4 100644
--- a/clang/test/SemaSYCL/spec-const-kernel-arg.cpp
+++ b/clang/test/SemaSYCL/spec-const-kernel-arg.cpp
@@ -1,4 +1,4 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s
+// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %s | FileCheck %s
 
 // This test checks that compiler generates correct initialization for spec
 // constants
@@ -6,24 +6,24 @@
 #include "Inputs/sycl.hpp"
 
 struct SpecConstantsWrapper {
-  cl::sycl::ONEAPI::experimental::spec_constant<int, class sc_name1> SC1;
-  cl::sycl::ONEAPI::experimental::spec_constant<int, class sc_name2> SC2;
+  sycl::ONEAPI::experimental::spec_constant<int, class sc_name1> SC1;
+  sycl::ONEAPI::experimental::spec_constant<int, class sc_name2> SC2;
 };
 
 int main() {
-  cl::sycl::ONEAPI::experimental::spec_constant<char, class MyInt32Const> SC;
-  SpecConstantsWrapper W;
-  cl::sycl::kernel_single_task<class kernel_sc>(
+  sycl::ONEAPI::experimental::spec_constant<char, class MyInt32Const> SC;
+  SpecConstantsWrapper SCWrapper;
+  sycl::kernel_single_task<class kernel_sc>(
       [=]() {
         (void)SC;
-        (void)W;
+        (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<char, class MyInt32Const>':'cl::sycl::ONEAPI::experimental::spec_constant<char, MyInt32Const>'
+// CHECK-NEXT: CXXConstructExpr {{.*}}'sycl::ONEAPI::experimental::spec_constant<char, class MyInt32Const>':'sycl::ONEAPI::experimental::spec_constant<char, MyInt32Const>'
 // CHECK-NEXT: InitListExpr {{.*}} 'SpecConstantsWrapper'
-// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::ONEAPI::experimental::spec_constant<int, class sc_name1>':'cl::sycl::ONEAPI::experimental::spec_constant<int, sc_name1>'
-// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::ONEAPI::experimental::spec_constant<int, class sc_name2>':'cl::sycl::ONEAPI::experimental::spec_constant<int, sc_name2>'
+// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::ONEAPI::experimental::spec_constant<int, class sc_name1>':'sycl::ONEAPI::experimental::spec_constant<int, sc_name1>'
+// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::ONEAPI::experimental::spec_constant<int, class sc_name2>':'sycl::ONEAPI::experimental::spec_constant<int, sc_name2>'

From aca853de7b709661a5052879e2b8d2855b626079 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Mon, 11 Jan 2021 18:43:12 -0800
Subject: [PATCH 13/22] Update sampler.cpp

---
 clang/test/SemaSYCL/sampler.cpp               | 24 +++++++++++--------
 clang/test/SemaSYCL/spec-const-kernel-arg.cpp | 14 +++++++----
 2 files changed, 23 insertions(+), 15 deletions(-)

diff --git a/clang/test/SemaSYCL/sampler.cpp b/clang/test/SemaSYCL/sampler.cpp
index 461615fa3fa9a..0420028522cc8 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 -triple spir64 -ast-dump -sycl-std=2020 %s | FileCheck %s
+
+// This test checks if the compiler correctly initilaizes the SYCL Sampler object when passed as a kernel argument.
 
 #include "Inputs/sycl.hpp"
 
-template <typename name, typename Func>
-__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) {
-  kernelFunc();
-}
+sycl::queue myQueue;
 
 int main() {
-  cl::sycl::sampler Sampler;
-  kernel<class use_kernel_for_test>([=]() {
-    Sampler.use();
+
+  sycl::sampler Sampler;
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class SamplerLambda>([=] {
+      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 240241062ccb4..7027d6bc016ec 100644
--- a/clang/test/SemaSYCL/spec-const-kernel-arg.cpp
+++ b/clang/test/SemaSYCL/spec-const-kernel-arg.cpp
@@ -5,6 +5,8 @@
 
 #include "Inputs/sycl.hpp"
 
+sycl::queue myQueue;
+
 struct SpecConstantsWrapper {
   sycl::ONEAPI::experimental::spec_constant<int, class sc_name1> SC1;
   sycl::ONEAPI::experimental::spec_constant<int, class sc_name2> SC2;
@@ -13,11 +15,13 @@ struct SpecConstantsWrapper {
 int main() {
   sycl::ONEAPI::experimental::spec_constant<char, class MyInt32Const> SC;
   SpecConstantsWrapper SCWrapper;
-  sycl::kernel_single_task<class kernel_sc>(
-      [=]() {
-        (void)SC;
-        (void)SCWrapper;
-      });
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class kernel_sc>(
+        [=] {
+          (void)SC;
+          (void)SCWrapper;
+        });
+  });
 }
 
 // CHECK: FunctionDecl {{.*}}kernel_sc{{.*}} 'void ()'

From d1cbee305303536bba3043cfbda35f4c44d69e87 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Tue, 12 Jan 2021 12:19:40 -0800
Subject: [PATCH 14/22] Update streams.cpp

---
 clang/test/SemaSYCL/streams.cpp | 751 ++++++++++++++++----------------
 1 file changed, 377 insertions(+), 374 deletions(-)

diff --git a/clang/test/SemaSYCL/streams.cpp b/clang/test/SemaSYCL/streams.cpp
index 15c1c8625a7ad..17425e741192f 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 -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 <typename Name, typename Func>
-__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) {
-  kernelFunc();
-}
+#include "Inputs/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,91 @@ 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<class stream_test>([=]() {
-    in_lambda.use();
-    in_lambda_array[1].use();
-    in_lambda_mdarray[1][1].use();
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class stream_test>([=]() {
+      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 <NoOp>
-// CHECK-NEXT: DeclRefExpr {{.*}} 'cl::sycl::stream' lvalue ParmVar
+
+// CHECK: InitListExpr {{.*}}
+// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept'
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar
 
 // HasArrayOfHasStreams
 // CHECK-NEXT: InitListExpr {{.*}} 'HasArrayOfHasStreams'
@@ -113,35 +116,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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar
 
 // HasArrayOfHasStreams Array
 // CHECK-NEXT: InitListExpr {{.*}} 'HasArrayOfHasStreams [2]'
@@ -154,35 +157,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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar
 // CHECK-NEXT: InitListExpr {{.*}} 'HasArrayOfHasStreams'
 // HasArrayOfHasStreams::i
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' <LValueToRValue>
@@ -192,222 +195,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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 +418,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -426,7 +429,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -435,11 +438,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -449,9 +452,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -461,11 +464,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -475,9 +478,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -487,9 +490,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -498,7 +501,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -507,11 +510,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -521,9 +524,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -533,11 +536,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -547,9 +550,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -561,9 +564,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -575,7 +578,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -587,11 +590,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -604,9 +607,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -619,11 +622,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -636,9 +639,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -651,9 +654,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -665,7 +668,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -677,11 +680,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -694,9 +697,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -709,11 +712,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -726,9 +729,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -741,9 +744,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -755,7 +758,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -767,11 +770,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -784,9 +787,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -799,11 +802,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -816,9 +819,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -831,9 +834,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -845,7 +848,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -857,11 +860,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -874,9 +877,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -889,11 +892,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -906,9 +909,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -924,24 +927,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue
 // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at
 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1
 
@@ -949,44 +952,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'cl::sycl::stream [2]' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'cl::sycl::stream (*)[2]' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2][2]' lvalue
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream [2]' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream (*)[2]' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2][2]' lvalue
 // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at
 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1
 // CHECK-NEXT: IntegerLiteral {{.*}} '{{.*}}' 1
@@ -994,24 +997,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// 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 +1023,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1030,9 +1033,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1043,9 +1046,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1056,7 +1059,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1066,9 +1069,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1079,9 +1082,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1094,7 +1097,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1107,9 +1110,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1123,9 +1126,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1139,7 +1142,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1152,9 +1155,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1168,9 +1171,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1184,7 +1187,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1197,9 +1200,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1213,9 +1216,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1229,7 +1232,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 *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1242,9 +1245,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs
@@ -1258,9 +1261,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 *' <ArrayToPointerDecay>
-// CHECK-NEXT: MemberExpr {{.*}} 'cl::sycl::stream [2]' lvalue .s_array
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'sycl::stream *' <ArrayToPointerDecay>
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::stream [2]' lvalue .s_array
 // CHECK-NEXT: ArraySubscriptExpr {{.*}} 'HasStreams' lvalue
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'HasStreams *' <ArrayToPointerDecay>
 // CHECK-NEXT: MemberExpr {{.*}} 'HasStreams [2]' lvalue .hs

From cf36c62d2b619ab76f91d63c9d1d566a92d1aba1 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Tue, 12 Jan 2021 17:06:12 -0800
Subject: [PATCH 15/22] Update accessor_inheritance.cpp

---
 clang/test/SemaSYCL/accessor_inheritance.cpp | 56 +++++++++++---------
 1 file changed, 32 insertions(+), 24 deletions(-)

diff --git a/clang/test/SemaSYCL/accessor_inheritance.cpp b/clang/test/SemaSYCL/accessor_inheritance.cpp
index 57b3179fb997f..f44cfa2816423 100644
--- a/clang/test/SemaSYCL/accessor_inheritance.cpp
+++ b/clang/test/SemaSYCL/accessor_inheritance.cpp
@@ -1,49 +1,57 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump %s | FileCheck %s
+// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %s | FileCheck %s
+
+// This test checks inheritance support for struct types with accessors
+// passed as kernel arguments, which are decomposed to individual fields.
+
 #include "Inputs/sycl.hpp"
 
-struct Base {
+sycl::queue myQueue;
+
+struct AccessorBase {
   int A, B;
-  cl::sycl::accessor<char, 1, cl::sycl::access::mode::read> AccField;
+  sycl::accessor<char, 1, sycl::access::mode::read> AccField;
 };
 
-struct Captured : Base,
-                  cl::sycl::accessor<char, 1, cl::sycl::access::mode::read> {
+struct AccessorDerived : AccessorBase,
+                         sycl::accessor<char, 1, sycl::access::mode::read> {
   int C;
 };
 
 int main() {
-  Captured Obj;
-  cl::sycl::kernel_single_task<class kernel>(
-      [=]() {
-        Obj.use();
-      });
+  AccessorDerived DerivedObject;
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class kernel>(
+        [=]() {
+          DerivedObject.use();
+        });
+  });
 }
 
 // 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' <LValueToRValue>
 // CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_A' 'int'
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' <LValueToRValue>
 // CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_B' 'int'
-// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::accessor<char, 1, cl::sycl::access::mode::read>':'cl::sycl::accessor<char, 1, cl::sycl::access::mode::read, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>' 'void () noexcept'
-// CHECK-NEXT: CXXConstructExpr {{.*}} 'cl::sycl::accessor<char, 1, cl::sycl::access::mode::read>':'cl::sycl::accessor<char, 1, cl::sycl::access::mode::read, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>' 'void () noexcept'
+// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::accessor<char, 1, sycl::access::mode::read>':'sycl::accessor<char, 1, sycl::access::mode::read, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>' 'void () noexcept'
+// CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::accessor<char, 1, sycl::access::mode::read>':'sycl::accessor<char, 1, sycl::access::mode::read, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>' 'void () noexcept'
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' <LValueToRValue>
 // CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_C' 'int'
 
@@ -51,15 +59,15 @@ int main() {
 // CHECK: CXXMemberCallExpr {{.*}} 'void'
 // CHECK-NEXT: MemberExpr {{.*}} .__init
 // CHECK-NEXT: MemberExpr {{.*}} .AccField
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'Base' lvalue <DerivedToBase (Base)>
-// CHECK-NEXT: MemberExpr {{.*}} 'Captured' lvalue .
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'AccessorBase' lvalue <DerivedToBase (AccessorBase)>
+// CHECK-NEXT: MemberExpr {{.*}} 'AccessorDerived' lvalue .
 // CHECK-NEXT: DeclRefExpr {{.*}}'(lambda at {{.*}}accessor_inheritance.cpp
 // CHECK-NEXT: ImplicitCastExpr {{.*}} '__global char *' <LValueToRValue>
 // 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 *' <LValueToRValue>
 // CHECK-NEXT: DeclRefExpr {{.*}} '__global char *' lvalue ParmVar {{.*}} '_arg__base' '__global char *'

From a35dbb81de544181d4b5e01ad7275c73b58545ad Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Tue, 12 Jan 2021 17:13:51 -0800
Subject: [PATCH 16/22] Update wrapped-accessor.cpp

---
 clang/test/SemaSYCL/accessor_inheritance.cpp |  4 +-
 clang/test/SemaSYCL/wrapped-accessor.cpp     | 58 ++++++++++----------
 2 files changed, 33 insertions(+), 29 deletions(-)

diff --git a/clang/test/SemaSYCL/accessor_inheritance.cpp b/clang/test/SemaSYCL/accessor_inheritance.cpp
index f44cfa2816423..f0a7fbbb355f4 100644
--- a/clang/test/SemaSYCL/accessor_inheritance.cpp
+++ b/clang/test/SemaSYCL/accessor_inheritance.cpp
@@ -21,10 +21,12 @@ int main() {
   AccessorDerived DerivedObject;
   myQueue.submit([&](sycl::handler &h) {
     h.single_task<class kernel>(
-        [=]() {
+        [=] {
           DerivedObject.use();
         });
   });
+
+  return 0;
 }
 
 // Check kernel parameters
diff --git a/clang/test/SemaSYCL/wrapped-accessor.cpp b/clang/test/SemaSYCL/wrapped-accessor.cpp
index f087621e3e575..569e3f5b6f515 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 -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"
 
+sycl::queue myQueue;
+
 template <typename Acc>
 struct AccWrapper { Acc accessor; };
 
-template <typename name, typename Func>
-__attribute__((sycl_kernel)) void kernel(const Func &kernelFunc) {
-  kernelFunc();
-}
-
 int main() {
-  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write> acc;
+  sycl::accessor<int, 1, sycl::access::mode::read_write> acc;
   auto acc_wrapped = AccWrapper<decltype(acc)>{acc};
-  kernel<class wrapped_access>(
-      [=]() {
-        acc_wrapped.accessor.use();
-      });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class wrapped_access>(
+        [=] {
+          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<cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>>'
-// CHECK-NEXT: CXXConstructExpr {{.*}}'cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>':'cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>' 'void () noexcept'
+// CHECK-NEXT: InitListExpr {{.*}}'AccWrapper<sycl::accessor<int, 1, sycl::access::mode::read_write, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>>'
+// CHECK-NEXT: CXXConstructExpr {{.*}}'sycl::accessor<int, 1, sycl::access::mode::read_write, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>':'sycl::accessor<int, 1, sycl::access::mode::read_write, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>' '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<int, 1, cl::sycl::access::mode::read_write, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>':'cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>' lvalue .accessor {{.*}}
-// CHECK-NEXT: MemberExpr {{.*}} 'AccWrapper<cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>>':'AccWrapper<cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write, cl::sycl::access::target::global_buffer, cl::sycl::access::placeholder::false_t>>' lvalue .
+// CHECK-NEXT: MemberExpr {{.*}} 'sycl::accessor<int, 1, sycl::access::mode::read_write, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>':'sycl::accessor<int, 1, sycl::access::mode::read_write, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>' lvalue .accessor {{.*}}
+// CHECK-NEXT: MemberExpr {{.*}} 'AccWrapper<sycl::accessor<int, 1, sycl::access::mode::read_write, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>>':'AccWrapper<sycl::accessor<int, 1, sycl::access::mode::read_write, sycl::access::target::global_buffer, sycl::access::placeholder::false_t>>' lvalue .
 // CHECK-NEXT: DeclRefExpr {{.*}} '(lambda at {{.*}}wrapped-accessor.cpp{{.*}})' lvalue Var {{.*}} '(lambda at {{.*}}wrapped-accessor.cpp{{.*}})'
 
 // Parameters of the _init method
 // CHECK-NEXT: ImplicitCastExpr {{.*}} <LValueToRValue>
 // 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// 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 <NoOp>
-// 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 <NoOp>
+// CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::id<1>' lvalue ParmVar {{.*}} '[[_arg_Offset]]' 'sycl::id<1>'

From 2ccb46a559e2015f02e4e792316b70d447892e14 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Tue, 12 Jan 2021 18:14:00 -0800
Subject: [PATCH 17/22] Update array-kernel-param.cpp

---
 clang/test/SemaSYCL/array-kernel-param.cpp | 352 +++++++++++----------
 1 file changed, 183 insertions(+), 169 deletions(-)

diff --git a/clang/test/SemaSYCL/array-kernel-param.cpp b/clang/test/SemaSYCL/array-kernel-param.cpp
index 58eee1c4debef..970190081d024 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 -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"
 
-using namespace cl::sycl;
+sycl::queue myQueue;
 
-template <typename name, typename Func>
-__attribute__((sycl_kernel)) void a_kernel(const Func &kernelFunc) {
-  kernelFunc();
-}
+using namespace sycl;
 
 template <typename T>
 struct S {
@@ -20,102 +17,119 @@ struct S {
 int main() {
 
   using Accessor =
-      accessor<int, 1, access::mode::read_write, access::target::global_buffer>;
+      sycl::accessor<int, 1, sycl::access::mode::read_write, sycl::access::target::global_buffer>;
 
-  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<int> 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<class kernel_A>(
-      [=]() {
-        acc[1].use();
-      });
-
-  a_kernel<class kernel_B>(
-      [=]() {
-        int local = a[1];
-      });
-
-  a_kernel<class kernel_B_ptrs>(
-      [=]() {
-        int local = *a_ptrs[1];
-      });
-
-  a_kernel<class kernel_C>(
-      [=]() {
-        struct_acc.member_acc[2].use();
-      });
-
-  a_kernel<class kernel_D>(
-      [=]() {
-        foo local = struct_array[1];
-      });
-
-  a_kernel<class kernel_E>(
-      [=]() {
-        int local = s.a[2];
-      });
-
-  a_kernel<class kernel_F>(
-      [=]() {
-        int local = array_2D[1][1];
-      });
-
-  a_kernel<class kernel_G>(
-      [=]() {
-        foo2 local = struct_array2[0];
-      });
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class Kernel_Accessor>(
+        [=] {
+          ReadWriteAccessor[1].use();
+        });
+  });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class Kernel_Array>(
+        [=] {
+          int local = Array[1];
+        });
+  });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class Kernel_Array_Ptrs>(
+        [=] {
+          int local = *ArrayOfPointers[1];
+        });
+  });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class Kernel_StructAccArray>(
+        [=] {
+          StructAccArrayObj.member_acc[2].use();
+        });
+  });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class Kernel_DecomposedStruct>(
+        [=] {
+          DecomposedStruct local = DecompStructArray[1];
+        });
+  });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class Kernel_TemplatedStructArray>(
+        [=] {
+          int local = s.a[2];
+        });
+  });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class Kernel_Array_2D>(
+        [=] {
+          int local = array_2D[1][1];
+        });
+  });
+
+  myQueue.submit([&](sycl::handler &h) {
+    h.single_task<class Kernel_NonDecomposedStruct>(
+        [=] {
+          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 *' <LValueToRValue>
 // 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' <LValueToRValue>
-// 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' <LValueToRValue>
-// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_x' 'int'
+// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_x' 'int'
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' <LValueToRValue>
-// 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 *' <LValueToRValue>
 // 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 *' <LValueToRValue>
 // 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 *' <LValueToRValue>
 // 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 *' <LValueToRValue>
 // 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' <LValueToRValue>
-// 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' <LValueToRValue>
-// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_foo_inner_x' 'int'
+// CHECK-NEXT: DeclRefExpr {{.*}} 'int' lvalue ParmVar {{.*}} '_arg_x' 'int'
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'int' <LValueToRValue>
-// 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 *' <LValueToRValue>
 // 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 *' <LValueToRValue>
 // 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 *' <LValueToRValue>
 // 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 *' <LValueToRValue>
 // 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<int>)'
+// Check Kernel_TemplatedStructArray parameters
+// CHECK: FunctionDecl {{.*}}Kernel_TemplatedStructArray{{.*}} 'void (S<int>)'
 // CHECK-NEXT: ParmVarDecl {{.*}} used _arg_ 'S<int>':'S<int>'
 // CHECK-NEXT: CompoundStmt
 // CHECK-NEXT: DeclStmt
@@ -318,10 +332,10 @@ int main() {
 // CHECK-NEXT: ImplicitCastExpr
 // CHECK-NEXT: DeclRefExpr {{.*}} 'S<int>':'S<int>' lvalue ParmVar {{.*}} '_arg_' 'S<int>':'S<int>'
 
-// 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 <NoOp>
-// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'foo2' lvalue
-// CHECK-NEXT: ImplicitCastExpr {{.*}} 'foo2 *' <ArrayToPointerDecay>
-// 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 <NoOp>
+// CHECK-NEXT: ArraySubscriptExpr {{.*}} 'NonDecomposedStruct' lvalue
+// CHECK-NEXT: ImplicitCastExpr {{.*}} 'NonDecomposedStruct *' <ArrayToPointerDecay>
+// 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

From 184df8ba06915690a0a580fc55eacd8ce313173e Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Mon, 25 Jan 2021 15:04:14 -0800
Subject: [PATCH 18/22] Include sycl runtime headers as system headers.

---
 clang/test/SemaSYCL/accessor_inheritance.cpp      | 4 ++--
 clang/test/SemaSYCL/accessors-targets-image.cpp   | 4 ++--
 clang/test/SemaSYCL/accessors-targets.cpp         | 4 ++--
 clang/test/SemaSYCL/allow-constexpr-recursion.cpp | 4 ++--
 clang/test/SemaSYCL/array-kernel-param-neg.cpp    | 4 ++--
 clang/test/SemaSYCL/array-kernel-param.cpp        | 4 ++--
 clang/test/SemaSYCL/basic-kernel-wrapper.cpp      | 4 ++--
 clang/test/SemaSYCL/decomposition.cpp             | 4 ++--
 clang/test/SemaSYCL/fake-accessors.cpp            | 4 ++--
 clang/test/SemaSYCL/half-kernel-arg.cpp           | 4 ++--
 clang/test/SemaSYCL/sampler.cpp                   | 4 ++--
 clang/test/SemaSYCL/spec-const-kernel-arg.cpp     | 4 ++--
 clang/test/SemaSYCL/streams.cpp                   | 4 ++--
 clang/test/SemaSYCL/wrapped-accessor.cpp          | 4 ++--
 14 files changed, 28 insertions(+), 28 deletions(-)

diff --git a/clang/test/SemaSYCL/accessor_inheritance.cpp b/clang/test/SemaSYCL/accessor_inheritance.cpp
index f0a7fbbb355f4..4d445de173799 100644
--- a/clang/test/SemaSYCL/accessor_inheritance.cpp
+++ b/clang/test/SemaSYCL/accessor_inheritance.cpp
@@ -1,9 +1,9 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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 inheritance support for struct types with accessors
 // passed as kernel arguments, which are decomposed to individual fields.
 
-#include "Inputs/sycl.hpp"
+#include "sycl.hpp"
 
 sycl::queue myQueue;
 
diff --git a/clang/test/SemaSYCL/accessors-targets-image.cpp b/clang/test/SemaSYCL/accessors-targets-image.cpp
index 53888ad55930d..b7a516420650d 100644
--- a/clang/test/SemaSYCL/accessors-targets-image.cpp
+++ b/clang/test/SemaSYCL/accessors-targets-image.cpp
@@ -1,9 +1,9 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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 if the compiler generates correct kernel wrapper arguments for
 // image accessors targets.
 
-#include "Inputs/sycl.hpp"
+#include "sycl.hpp"
 
 sycl::queue q;
 
diff --git a/clang/test/SemaSYCL/accessors-targets.cpp b/clang/test/SemaSYCL/accessors-targets.cpp
index 5d90adeb44700..9b05d2ce10298 100644
--- a/clang/test/SemaSYCL/accessors-targets.cpp
+++ b/clang/test/SemaSYCL/accessors-targets.cpp
@@ -1,9 +1,9 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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 the compiler generates correct kernel wrapper arguments for
 // different accessors targets.
 
-#include "Inputs/sycl.hpp"
+#include "sycl.hpp"
 
 sycl::queue q;
 
diff --git a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
index 5c5542a61c077..eb8f213d9cd4e 100644
--- a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
+++ b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
@@ -1,8 +1,8 @@
-// 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 -Wno-sycl-2017-compat -verify -fsyntax-only -std=c++20 -Werror=vla %s
 
 // This test verifies that a SYCL kernel executed on a device, cannot call a recursive function.
 
-#include "Inputs/sycl.hpp"
+#include "sycl.hpp"
 
 sycl::queue q;
 
diff --git a/clang/test/SemaSYCL/array-kernel-param-neg.cpp b/clang/test/SemaSYCL/array-kernel-param-neg.cpp
index c16c22081e16e..b0a5bade19e84 100755
--- a/clang/test/SemaSYCL/array-kernel-param-neg.cpp
+++ b/clang/test/SemaSYCL/array-kernel-param-neg.cpp
@@ -1,10 +1,10 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -fcxx-exceptions -sycl-std=2020 -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.
 
-#include "Inputs/sycl.hpp"
+#include "sycl.hpp"
 
 sycl::queue q;
 
diff --git a/clang/test/SemaSYCL/array-kernel-param.cpp b/clang/test/SemaSYCL/array-kernel-param.cpp
index 970190081d024..9701e331ad236 100644
--- a/clang/test/SemaSYCL/array-kernel-param.cpp
+++ b/clang/test/SemaSYCL/array-kernel-param.cpp
@@ -1,9 +1,9 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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"
 
 sycl::queue myQueue;
 
diff --git a/clang/test/SemaSYCL/basic-kernel-wrapper.cpp b/clang/test/SemaSYCL/basic-kernel-wrapper.cpp
index 6bdbf3dc6fe78..6176fa3098322 100644
--- a/clang/test/SemaSYCL/basic-kernel-wrapper.cpp
+++ b/clang/test/SemaSYCL/basic-kernel-wrapper.cpp
@@ -1,9 +1,9 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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"
 
 sycl::queue myQueue;
 
diff --git a/clang/test/SemaSYCL/decomposition.cpp b/clang/test/SemaSYCL/decomposition.cpp
index f871445057b00..46a8203b52dd3 100644
--- a/clang/test/SemaSYCL/decomposition.cpp
+++ b/clang/test/SemaSYCL/decomposition.cpp
@@ -1,10 +1,10 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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 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.
 
-#include "Inputs/sycl.hpp"
+#include "sycl.hpp"
 
 sycl::queue myQueue;
 
diff --git a/clang/test/SemaSYCL/fake-accessors.cpp b/clang/test/SemaSYCL/fake-accessors.cpp
index fd1cc26d0e24e..b59f83078533a 100644
--- a/clang/test/SemaSYCL/fake-accessors.cpp
+++ b/clang/test/SemaSYCL/fake-accessors.cpp
@@ -1,6 +1,6 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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"
+#include "sycl.hpp"
 
 sycl::queue myQueue;
 
diff --git a/clang/test/SemaSYCL/half-kernel-arg.cpp b/clang/test/SemaSYCL/half-kernel-arg.cpp
index b579bde840cca..77bbafc3eb2bc 100644
--- a/clang/test/SemaSYCL/half-kernel-arg.cpp
+++ b/clang/test/SemaSYCL/half-kernel-arg.cpp
@@ -1,9 +1,9 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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 sycl::half type inside the OpenCL kernel
 
-#include "Inputs/sycl.hpp"
+#include "sycl.hpp"
 
 sycl::queue myQueue;
 
diff --git a/clang/test/SemaSYCL/sampler.cpp b/clang/test/SemaSYCL/sampler.cpp
index 0420028522cc8..ff5aace314a00 100644
--- a/clang/test/SemaSYCL/sampler.cpp
+++ b/clang/test/SemaSYCL/sampler.cpp
@@ -1,8 +1,8 @@
-// RUN: %clang_cc1 -S -fsycl -fsycl-is-device -triple spir64 -ast-dump -sycl-std=2020 %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
 
 // This test checks if the compiler correctly initilaizes the SYCL Sampler object when passed as a kernel argument.
 
-#include "Inputs/sycl.hpp"
+#include "sycl.hpp"
 
 sycl::queue myQueue;
 
diff --git a/clang/test/SemaSYCL/spec-const-kernel-arg.cpp b/clang/test/SemaSYCL/spec-const-kernel-arg.cpp
index 7027d6bc016ec..a2bf4caa96c68 100644
--- a/clang/test/SemaSYCL/spec-const-kernel-arg.cpp
+++ b/clang/test/SemaSYCL/spec-const-kernel-arg.cpp
@@ -1,9 +1,9 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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;
 
diff --git a/clang/test/SemaSYCL/streams.cpp b/clang/test/SemaSYCL/streams.cpp
index 17425e741192f..42834716f0cfc 100644
--- a/clang/test/SemaSYCL/streams.cpp
+++ b/clang/test/SemaSYCL/streams.cpp
@@ -1,9 +1,9 @@
-// RUN: %clang_cc1 -S -fsycl -fsycl-is-device -triple spir64 -ast-dump -sycl-std=2020 %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
 
 // This test demonstrates passing of SYCL stream instances as kernel arguments and checks if the compiler generates
 // the correct ast-dump
 
-#include "Inputs/sycl.hpp"
+#include "sycl.hpp"
 
 sycl::queue myQueue;
 
diff --git a/clang/test/SemaSYCL/wrapped-accessor.cpp b/clang/test/SemaSYCL/wrapped-accessor.cpp
index 569e3f5b6f515..e5825fac768f8 100644
--- a/clang/test/SemaSYCL/wrapped-accessor.cpp
+++ b/clang/test/SemaSYCL/wrapped-accessor.cpp
@@ -1,9 +1,9 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -ast-dump -sycl-std=2020 %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;
 

From a63ea025ed8590fbfaffedb61d96344218c794df Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Mon, 25 Jan 2021 15:19:10 -0800
Subject: [PATCH 19/22] Address review comments

---
 clang/test/SemaSYCL/accessors-targets-image.cpp | 5 -----
 clang/test/SemaSYCL/decomposition.cpp           | 6 +++---
 clang/test/SemaSYCL/sampler.cpp                 | 2 +-
 clang/test/SemaSYCL/streams.cpp                 | 3 ++-
 4 files changed, 6 insertions(+), 10 deletions(-)

diff --git a/clang/test/SemaSYCL/accessors-targets-image.cpp b/clang/test/SemaSYCL/accessors-targets-image.cpp
index b7a516420650d..075ec7e144938 100644
--- a/clang/test/SemaSYCL/accessors-targets-image.cpp
+++ b/clang/test/SemaSYCL/accessors-targets-image.cpp
@@ -9,7 +9,6 @@ sycl::queue q;
 
 int main() {
 
-  // 1-dimensional accessor with Read-only access
   sycl::accessor<int, 1, sycl::access::mode::read,
                  sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc1d_read;
@@ -32,7 +31,6 @@ int main() {
         });
   });
 
-  // 3-dimensional accessor with Read-only access
   sycl::accessor<int, 3, sycl::access::mode::read,
                  sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc3d_read;
@@ -44,7 +42,6 @@ int main() {
         });
   });
 
-  // 1-dimensional accessor with Write-only access
   sycl::accessor<int, 1, sycl::access::mode::write,
                  sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc1d_write;
@@ -55,7 +52,6 @@ int main() {
         });
   });
 
-  // 2-dimensional accessor with Write-only access
   sycl::accessor<int, 2, sycl::access::mode::write,
                  sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc2d_write;
@@ -66,7 +62,6 @@ int main() {
         });
   });
 
-  // 3-dimensional accessor with Write-only access
   sycl::accessor<int, 3, sycl::access::mode::write,
                  sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc3d_write;
diff --git a/clang/test/SemaSYCL/decomposition.cpp b/clang/test/SemaSYCL/decomposition.cpp
index 46a8203b52dd3..83ac5a6cdc653 100644
--- a/clang/test/SemaSYCL/decomposition.cpp
+++ b/clang/test/SemaSYCL/decomposition.cpp
@@ -64,11 +64,11 @@ int main() {
 
   StructNonDecomposed SimpleStruct;
   StructNonDecomposed ArrayOfSimpleStruct[5];
-  StructWithNonDecomposedStruct d2;
-  StructWithNonDecomposedStruct d2s[5];
+  StructWithNonDecomposedStruct NonDecompStruct;
+  StructWithNonDecomposedStruct ArrayOfNonDecompStruct[5];
   // Check to ensure that these are not decomposed.
   myQueue.submit([&](sycl::handler &h) {
-    h.single_task<class NonDecomposed>([=]() { return SimpleStruct.i + ArrayOfSimpleStruct[0].i + d2.i + d2s[0].i; });
+    h.single_task<class NonDecomposed>([=]() { return SimpleStruct.i + ArrayOfSimpleStruct[0].i + NonDecompStruct.i + ArrayOfNonDecompStruct[0].i; });
   });
   // CHECK: FunctionDecl {{.*}}NonDecomposed{{.*}} 'void (StructNonDecomposed, __wrapper_class, StructWithNonDecomposedStruct, __wrapper_class)'
 
diff --git a/clang/test/SemaSYCL/sampler.cpp b/clang/test/SemaSYCL/sampler.cpp
index ff5aace314a00..032eec95c2c20 100644
--- a/clang/test/SemaSYCL/sampler.cpp
+++ b/clang/test/SemaSYCL/sampler.cpp
@@ -1,6 +1,6 @@
 // RUN: %clang_cc1 -S -fsycl -fsycl-is-device -internal-isystem %S/Inputs -triple spir64 -ast-dump -sycl-std=2020 %s | FileCheck %s
 
-// This test checks if the compiler correctly initilaizes the SYCL Sampler object when passed as a kernel argument.
+// This test checks if the compiler correctly initializes the SYCL Sampler object when passed as a kernel argument.
 
 #include "sycl.hpp"
 
diff --git a/clang/test/SemaSYCL/streams.cpp b/clang/test/SemaSYCL/streams.cpp
index 42834716f0cfc..bae0671bf1358 100644
--- a/clang/test/SemaSYCL/streams.cpp
+++ b/clang/test/SemaSYCL/streams.cpp
@@ -50,7 +50,8 @@ int main() {
 
 // Initializers:
 
-// CHECK: InitListExpr {{.*}}
+// CHECK: InitListExpr {{.*}} '(lambda at
+// 'in_lambda'
 // CHECK-NEXT: CXXConstructExpr {{.*}} 'sycl::stream':'sycl::stream' 'void (const sycl::stream &) noexcept'
 // CHECK-NEXT: ImplicitCastExpr {{.*}} 'const sycl::stream' lvalue <NoOp>
 // CHECK-NEXT: DeclRefExpr {{.*}} 'sycl::stream':'sycl::stream' lvalue ParmVar

From c06c36508616769842e159c860480b5d52fae1cf Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Mon, 25 Jan 2021 15:32:34 -0800
Subject: [PATCH 20/22] Add -syc-std=2020

---
 clang/test/SemaSYCL/allow-constexpr-recursion.cpp | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
index eb8f213d9cd4e..6ab66376291b2 100644
--- a/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
+++ b/clang/test/SemaSYCL/allow-constexpr-recursion.cpp
@@ -1,4 +1,4 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -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
 
 // This test verifies that a SYCL kernel executed on a device, cannot call a recursive function.
 

From d4942b85ec5520d707ee35f41076f7cc7a655e14 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Thu, 28 Jan 2021 00:07:45 -0800
Subject: [PATCH 21/22] Fix fake-accessors test

---
 clang/test/SemaSYCL/fake-accessors.cpp | 30 ++++++++++++++------------
 1 file changed, 16 insertions(+), 14 deletions(-)

diff --git a/clang/test/SemaSYCL/fake-accessors.cpp b/clang/test/SemaSYCL/fake-accessors.cpp
index b59f83078533a..374eb320b71ff 100644
--- a/clang/test/SemaSYCL/fake-accessors.cpp
+++ b/clang/test/SemaSYCL/fake-accessors.cpp
@@ -1,8 +1,8 @@
-// RUN: %clang_cc1 -fsycl -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %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 "sycl.hpp"
 
-sycl::queue myQueue;
+sycl::queue deviceQueue;
 
 namespace fake {
 namespace cl {
@@ -21,31 +21,33 @@ class accessor {
 };
 
 int main() {
+
   fake::cl::sycl::accessor FakeAccessor = {1};
-  accessor acc1 = {1};
+  accessor AccessorClass = {1};
+
+  typedef sycl::accessor<int, 1, sycl::access::mode::read_write, cl::sycl::access::target::global_buffer>
+      MyAccessorTD;
+  MyAccessorTD AccessorTypeDef;
+
+  using MyAccessorA = sycl::accessor<int, 1, sycl::access::mode::read_write, cl::sycl::access::target::global_buffer>;
+  MyAccessorA AccessorAlias;
 
-  sycl::accessor<int, 1, sycl::access::mode::read_write> accessorA;
-  sycl::accessor<int, 1, sycl::access::mode::read_write> accessorB;
-  sycl::accessor<int, 1, sycl::access::mode::read_write> accessorC;
+  cl::sycl::accessor<int, 1, cl::sycl::access::mode::read_write> AccessorRegular;
 
-  myQueue.submit([&](sycl::handler &h) {
+  deviceQueue.submit([&](sycl::handler &h) {
     h.single_task<class fake_accessors>(
         [=] {
-          accessorA.use((void *)(FakeAccessor.field + acc1.field));
+          AccessorRegular.use((void *)(FakeAccessor.field + AccessorClass.field));
         });
-  });
 
-  myQueue.submit([&](sycl::handler &h) {
     h.single_task<class accessor_typedef>(
         [=] {
-          accessorB.use((void *)(FakeAccessor.field + acc1.field));
+          AccessorTypeDef.use((void *)(FakeAccessor.field + AccessorClass.field));
         });
-  });
 
-  myQueue.submit([&](sycl::handler &h) {
     h.single_task<class accessor_alias>(
         [=] {
-          accessorC.use((void *)(FakeAccessor.field + acc1.field));
+          AccessorAlias.use((void *)(FakeAccessor.field + AccessorClass.field));
         });
   });
 

From 7a0a3bf9173fde5ea2685a3f1f732084e33e4c92 Mon Sep 17 00:00:00 2001
From: "srividya.sundaram" <srividya.sundaram@intel.com>
Date: Thu, 28 Jan 2021 13:40:09 -0800
Subject: [PATCH 22/22] Remove necessary comment

---
 clang/test/SemaSYCL/accessors-targets-image.cpp | 1 -
 1 file changed, 1 deletion(-)

diff --git a/clang/test/SemaSYCL/accessors-targets-image.cpp b/clang/test/SemaSYCL/accessors-targets-image.cpp
index 075ec7e144938..c6d5a19168a84 100644
--- a/clang/test/SemaSYCL/accessors-targets-image.cpp
+++ b/clang/test/SemaSYCL/accessors-targets-image.cpp
@@ -20,7 +20,6 @@ int main() {
         });
   });
 
-  // 2-dimensional accessor with Read-only access
   sycl::accessor<int, 2, sycl::access::mode::read,
                  sycl::access::target::image, sycl::access::placeholder::false_t>
       image_acc2d_read;