Add front end lit tests for host pipes

Author: rho180

clang/test/CodeGenSYCL/Inputs/sycl.hpp

@@ -151,6 +151,34 @@ class [[__sycl_detail__::device_global]] [[__sycl_detail__::global_variable_allo
 } // namespace oneapi
 } // namespace ext
 
+namespace ext {
+namespace intel {
+namespace experimental {
+
+// host_pipe class decorated with attribute
+template <class _name, class _dataT>
+class
+host_pipe {
+
+public:
+  struct
+#ifdef __SYCL_DEVICE_ONLY__
+  // [[ __sycl_detail__::add_ir_attributes_global_variable(
+    // "sycl-host-pipe", nullptr)]] [[__sycl_detail__::host_pipe]]
+  [[__sycl_detail__::host_pipe]]
+#endif
+  __pipeType { const char __p; };
+
+  static constexpr __pipeType __pipe = {0};
+  static _dataT read() {
+    (void)__pipe;
+  }
+};
+
+} // namespace experimental
+} // namespace intel
+} // namespace ext
+
 template <int dim>
 struct id {
   template <typename... T>

clang/test/CodeGenSYCL/host_pipe.cpp

@@ -0,0 +1,29 @@
+// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -fsycl-unique-prefix=THE_PREFIX -std=c++17 -opaque-pointers -emit-llvm %s -o - | FileCheck %s
+#include "sycl.hpp"
+
+// Test cases below show that 'sycl-unique-id' LLVM IR attribute is attached to the
+// global variable whose type is decorated with host_pipe attribute, and that a
+// unique string is generated.
+
+// XFAIL:*
+
+using namespace sycl::ext::intel::experimental;
+using namespace sycl;
+queue q;
+
+// check that "sycl-unique-id" attribute is created for host pipes
+// CHECK: @_ZN4sycl3_V13ext5intel12experimental9host_pipeIZZZ3foovENKUlRNS0_7handlerEE_clES6_ENKUlvE_clEvE5HPIntiE6__pipeE = internal addrspace(1) constant %"struct.sycl::_V1::ext::intel::experimental::host_pipe<HPInt, int>::__pipeType" zeroinitializer, align 1 #[[HPINT_ATTRS:[0-9]+]]
+// CHECK: @_ZN4sycl3_V13ext5intel12experimental9host_pipeIZZZ3foovENKUlRNS0_7handlerEE_clES6_ENKUlvE_clEvE7HPFloatiE6__pipeE = internal addrspace(1) constant %"struct.sycl::_V1::ext::intel::experimental::host_pipe<HPFloat, int>::__pipeType" zeroinitializer, align 1 #[[HPFLOAT_ATTRS:[0-9]+]]
+
+void foo() {
+  q.submit([&](handler &h) {
+    h.single_task<class kernel_name_1>([=]() {
+      host_pipe<class HPInt, int>::read();
+      host_pipe<class HPFloat, int>::read();
+    });
+  });
+}
+
+// CHECK: attributes #[[HPINT_ATTRS]] = { "sycl-unique-id"="THE_PREFIX____ZN4sycl3_V13ext5intel12experimental9host_pipeIZZZ3foovENKUlRNS0_7handlerEE_clES6_ENKUlvE_clEvE5HPIntiE6__pipeE" }
+// CHECK: attributes #[[HPFLOAT_ATTRS]] = { "sycl-unique-id"="THE_PREFIX____ZN4sycl3_V13ext5intel12experimental9host_pipeIZZZ3foovENKUlRNS0_7handlerEE_clES6_ENKUlvE_clEvE7HPFloatiE6__pipeE" 
+

clang/test/CodeGenSYCL/host_pipe_int_footer_header.cpp

@@ -0,0 +1,48 @@
+// RUN: %clang_cc1 -fsycl-is-device -std=c++17 -internal-isystem %S/Inputs -triple spir64-unknown-unknown -fsycl-int-footer=%t.footer.h -fsycl-int-header=%t.header.h -fsycl-unique-prefix=THE_PREFIX %s -emit-llvm -o %t.ll
+// RUN: FileCheck -input-file=%t.footer.h %s --check-prefix=CHECK-FOOTER
+// RUN: FileCheck -input-file=%t.header.h %s --check-prefix=CHECK-HEADER
+#include "sycl.hpp"
+
+// Test cases below show that 'sycl-unique-id' LLVM IR attribute is attached to the
+// global variable whose type is decorated with host_pipe attribute, and that a
+// unique string is generated.
+
+using namespace sycl::ext::intel::experimental;
+using namespace sycl;
+queue q;
+
+void foo() {
+  q.submit([&](handler &h) {
+    h.single_task<class kernel_name_1>([=]() {
+      host_pipe<class HPInt, int>::read();
+      host_pipe<class HPFloat, int>::read();
+    });
+  });
+}
+
+// CHECK-HEADER: namespace sycl {
+// CHECK-HEADER-NEXT: __SYCL_INLINE_VER_NAMESPACE(_V1) {
+// CHECK-HEADER-NEXT: namespace detail {
+// CHECK-HEADER-NEXT: namespace {
+// CHECK-HEADER-NEXT: class __sycl_host_pipe_registration {
+// CHECK-HEADER-NEXT: public:
+// CHECK-HEADER-NEXT: __sycl_host_pipe_registration() noexcept;
+// CHECK-HEADER-NEXT: };
+// CHECK-HEADER-NEXT: __sycl_host_pipe_registration __sycl_host_pipe_registrar;
+// CHECK-HEADER-NEXT: } // namespace
+// CHECK-HEADER: } // namespace detail
+// CHECK-HEADER: } // __SYCL_INLINE_VER_NAMESPACE(_V1)
+// CHECK-HEADER: } // namespace sycl
+
+// CHECK-FOOTER: #include <sycl/detail/defines_elementary.hpp>
+// CHECK-FOOTER: #include <sycl/detail/host_pipe_map.hpp>
+// CHECK-FOOTER-NEXT: namespace sycl::detail {
+// CHECK-FOOTER-NEXT: namespace {
+// CHECK-FOOTER-NEXT: __sycl_host_pipe_registration::__sycl_host_pipe_registration() noexcept {
+
+// CHECK-FOOTER: host_pipe_map::add((void *)&::sycl::ext::intel::experimental::host_pipe<HPInt, int>::__pipe, "THE_PREFIX____ZN4sycl3_V13ext5intel12experimental9host_pipeIZZZ3foovENKUlRNS0_7handlerEE_clES6_ENKUlvE_clEvE5HPIntiE6__pipeE");
+// CHECK-FOOTER: host_pipe_map::add((void *)&::sycl::ext::intel::experimental::host_pipe<HPFloat, int>::__pipe, "THE_PREFIX____ZN4sycl3_V13ext5intel12experimental9host_pipeIZZZ3foovENKUlRNS0_7handlerEE_clES6_ENKUlvE_clEvE7HPFloatiE6__pipeE");
+
+// CHECK-FOOTER: } // namespace (unnamed)
+// CHECK-FOOTER: } // namespace sycl::detail
+