From 8dda2184b0127ef2cd6969ebb9d967a9b62ff3bc Mon Sep 17 00:00:00 2001 From: Alexander Batashev Date: Wed, 23 Sep 2020 13:57:28 +0300 Subject: [PATCH 1/4] [SYCL][NFC] Extend ABI test suite Add tests that check mangling for user code, when SYCL objects are part of function signature. --- sycl/test/abi/user_mangling.cpp | 87 +++++++++++++++++++++++++++++++++ 1 file changed, 87 insertions(+) create mode 100644 sycl/test/abi/user_mangling.cpp diff --git a/sycl/test/abi/user_mangling.cpp b/sycl/test/abi/user_mangling.cpp new file mode 100644 index 0000000000000..0ff73b545ca8b --- /dev/null +++ b/sycl/test/abi/user_mangling.cpp @@ -0,0 +1,87 @@ +// RUN: %clangxx -fsycl -c -emit-llvm -S -o - %s | FileCheck %s --check-prefix CHK-HOST +// RUN: %clangxx -fsycl -fsycl-device-only -O0 -c -emit-llvm -S -o - %s | FileCheck %s --check-prefix CHK-DEVICE + +#include + +#ifdef __SYCL_DEVICE_ONLY__ +// CHK-DEVICE: define dso_local spir_func void @_Z3accN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2014ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE(%"class._ZTSN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2014ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE.cl::sycl::accessor"* byval(%"class._ZTSN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2014ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE.cl::sycl::accessor") align 8 %0) +SYCL_EXTERNAL void acc(sycl::accessor) {} + +// CHK-DEVICE: define dso_local spir_func void @_Z11private_memN2cl4sycl14private_memoryIiLi1EEE(%"class._ZTSN2cl4sycl14private_memoryIiLi1EEE.cl::sycl::private_memory"* byval(%"class._ZTSN2cl4sycl14private_memoryIiLi1EEE.cl::sycl::private_memory") align 4 %0) +SYCL_EXTERNAL void private_mem(sycl::private_memory) {}; + +// CHK-DEVICE: define dso_local spir_func void @_Z5rangeN2cl4sycl5rangeILi1EEE(%"class._ZTSN2cl4sycl5rangeILi1EEE.cl::sycl::range"* byval(%"class._ZTSN2cl4sycl5rangeILi1EEE.cl::sycl::range") align 8 %0) +SYCL_EXTERNAL void range(sycl::range<1>) {} + +// CHK-DEVICE: define dso_local spir_func void @_Z2idN2cl4sycl2idILi1EEE(%"class._ZTSN2cl4sycl2idILi1EEE.cl::sycl::id"* byval(%"class._ZTSN2cl4sycl2idILi1EEE.cl::sycl::id") align 8 %0) +SYCL_EXTERNAL void id(sycl::id<1>) {} + +// CHK-DEVICE: define dso_local spir_func void @_Z4itemN2cl4sycl2idILi1EEE(%"class._ZTSN2cl4sycl2idILi1EEE.cl::sycl::id"* byval(%"class._ZTSN2cl4sycl2idILi1EEE.cl::sycl::id") align 8 %0) +SYCL_EXTERNAL void item(sycl::id<1>) {} + +// CHK-DEVICE: define dso_local spir_func void @_Z3vecN2cl4sycl3vecIiLi16EEE(%"class._ZTSN2cl4sycl3vecIiLi16EEE.cl::sycl::vec"* byval(%"class._ZTSN2cl4sycl3vecIiLi16EEE.cl::sycl::vec") align 64 %0) +SYCL_EXTERNAL void vec(sycl::vec) {} + +// CHK-DEVICE: define dso_local spir_func void @_Z6streamN2cl4sycl6streamE(%"class._ZTSN2cl4sycl6streamE.cl::sycl::stream"* byval(%"class._ZTSN2cl4sycl6streamE.cl::sycl::stream") align 8 %0) +SYCL_EXTERNAL void stream(sycl::stream) {} + +// CHK-DEVICE: define dso_local spir_func void @_Z7samplerN2cl4sycl7samplerE(%"class._ZTSN2cl4sycl7samplerE.cl::sycl::sampler"* byval(%"class._ZTSN2cl4sycl7samplerE.cl::sycl::sampler") align 8 %0) +SYCL_EXTERNAL void sampler(sycl::sampler) {} +#else +// CHK-HOST: define dso_local void @_Z3accN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2018ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE(%"class.cl::sycl::accessor"* %0) +void acc(sycl::accessor) {} + +// CHK-HOST: define dso_local void @_Z3bufN2cl4sycl6bufferIiLi1ENS0_6detail17aligned_allocatorIcEEvEE(%"class.cl::sycl::buffer"* %0) +void buf(sycl::buffer) {} + +// CHK-HOST: define dso_local void @_Z3ctxN2cl4sycl7contextE(%"class.cl::sycl::context"* %0) +void ctx(sycl::context) {} + +// CHK-HOST: define dso_local void @_Z6deviceN2cl4sycl6deviceE(%"class.cl::sycl::device"* %0) +void device(sycl::device) {} + +// CHK-HOST: define dso_local void @_Z10device_evtN2cl4sycl12device_eventE(i8** %.coerce) +void device_evt(sycl::device_event) {} + +// CHK-HOST: define dso_local void @_Z5eventN2cl4sycl5eventE(%"class.cl::sycl::event"* %0) +void event(sycl::event) {} + +// CHK-HOST: define dso_local void @_Z15device_selectorRN2cl4sycl15device_selectorE(%"class.cl::sycl::device_selector"* nonnull align 8 dereferenceable(8) %0) +void device_selector(sycl::device_selector&) {} + +// CHK-HOST: define dso_local void @_Z7handlerRN2cl4sycl7handlerE(%"class.cl::sycl::handler"* nonnull align 8 dereferenceable(560) %0) +void handler(sycl::handler&) {} + +// CHK-HOST: define dso_local void @_Z5imageN2cl4sycl5imageILi1ENS0_6detail17aligned_allocatorIhEEEE(%"class.cl::sycl::image"* %0) +void image(sycl::image<1>) {} + +// CHK-HOST: define dso_local void @_Z5rangeN2cl4sycl5rangeILi1EEE(i64 %.coerce) +void range(sycl::range<1>) {} + +// CHK-HOST: define dso_local void @_Z2idN2cl4sycl2idILi1EEE(i64 %.coerce) +void id(sycl::id<1>) {} + +// CHK-HOST: define dso_local void @_Z4itemN2cl4sycl4itemILi1ELb1EEE(%"class.cl::sycl::item"* byval(%"class.cl::sycl::item") align 8 %0) +void item(sycl::item<1>) {} + +// CHK-HOST: define dso_local void @_Z6streamN2cl4sycl6streamE(%"class.cl::sycl::stream"* %0) +void stream(sycl::stream) {} + +// CHK-HOST: define dso_local void @_Z7samplerN2cl4sycl7samplerE(%"class.cl::sycl::sampler"* %0) +void sampler(sycl::sampler) {} + +// CHK-HOST: define dso_local void @_Z5queueN2cl4sycl5queueE(%"class.cl::sycl::queue"* %0) +void queue(sycl::queue) {} + +// CHK-HOST: define dso_local void @_Z7programN2cl4sycl7programE(%"class.cl::sycl::program"* %0) +void program(sycl::program) {} + +// CHK-HOST: define dso_local void @_Z6kernelN2cl4sycl6kernelE(%"class.cl::sycl::kernel"* %0) +void kernel(sycl::kernel) {} + +// CHK-HOST: define dso_local void @_Z8platformN2cl4sycl8platformE(%"class.cl::sycl::platform"* %0) +void platform(sycl::platform) {} + +// CHK-HOST: define dso_local void @_Z3vecN2cl4sycl3vecIiLi16EEE(%"class.cl::sycl::vec"* %0) +void vec(sycl::vec) {} +#endif From 293686b01974cb28285ada71af3299ae5fd9910c Mon Sep 17 00:00:00 2001 From: Alexander Batashev Date: Wed, 23 Sep 2020 14:02:15 +0300 Subject: [PATCH 2/4] Disable clang-format as it breaks FileCheck directives --- sycl/test/abi/user_mangling.cpp | 1 + 1 file changed, 1 insertion(+) diff --git a/sycl/test/abi/user_mangling.cpp b/sycl/test/abi/user_mangling.cpp index 0ff73b545ca8b..3881cb9aab972 100644 --- a/sycl/test/abi/user_mangling.cpp +++ b/sycl/test/abi/user_mangling.cpp @@ -1,3 +1,4 @@ +// clang-format off // RUN: %clangxx -fsycl -c -emit-llvm -S -o - %s | FileCheck %s --check-prefix CHK-HOST // RUN: %clangxx -fsycl -fsycl-device-only -O0 -c -emit-llvm -S -o - %s | FileCheck %s --check-prefix CHK-DEVICE From 9595083868f37a4052fcd6d0d707fd8a2c6b4c29 Mon Sep 17 00:00:00 2001 From: Alexander Batashev Date: Wed, 23 Sep 2020 14:46:00 +0300 Subject: [PATCH 3/4] Add more tests --- sycl/test/abi/user_mangling.cpp | 12 ++++++++++++ 1 file changed, 12 insertions(+) diff --git a/sycl/test/abi/user_mangling.cpp b/sycl/test/abi/user_mangling.cpp index 3881cb9aab972..d096ee27fa347 100644 --- a/sycl/test/abi/user_mangling.cpp +++ b/sycl/test/abi/user_mangling.cpp @@ -8,6 +8,12 @@ // CHK-DEVICE: define dso_local spir_func void @_Z3accN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2014ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE(%"class._ZTSN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2014ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE.cl::sycl::accessor"* byval(%"class._ZTSN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2014ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE.cl::sycl::accessor") align 8 %0) SYCL_EXTERNAL void acc(sycl::accessor) {} +// CHK-DEVICE: define dso_local spir_func void @_Z3accN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2016ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE(%"class._ZTSN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2016ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE.cl::sycl::accessor"* byval(%"class._ZTSN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2016ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE.cl::sycl::accessor") align 8 %0) +SYCL_EXTERNAL void acc(sycl::accessor) {} + +// CHK-DEVICE: define dso_local spir_func void @_Z3accN2cl4sycl8accessorINS0_3vecIiLi4EEELi1ELNS0_6access4modeE1024ELNS4_6targetE2017ELNS4_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE(%"class._ZTSN2cl4sycl8accessorINS0_3vecIiLi4EEELi1ELNS0_6access4modeE1024ELNS4_6targetE2017ELNS4_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE.cl::sycl::accessor"* byval(%"class._ZTSN2cl4sycl8accessorINS0_3vecIiLi4EEELi1ELNS0_6access4modeE1024ELNS4_6targetE2017ELNS4_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE.cl::sycl::accessor") align 8 %0) +SYCL_EXTERNAL void acc(sycl::accessor) {} + // CHK-DEVICE: define dso_local spir_func void @_Z11private_memN2cl4sycl14private_memoryIiLi1EEE(%"class._ZTSN2cl4sycl14private_memoryIiLi1EEE.cl::sycl::private_memory"* byval(%"class._ZTSN2cl4sycl14private_memoryIiLi1EEE.cl::sycl::private_memory") align 4 %0) SYCL_EXTERNAL void private_mem(sycl::private_memory) {}; @@ -32,6 +38,12 @@ SYCL_EXTERNAL void sampler(sycl::sampler) {} // CHK-HOST: define dso_local void @_Z3accN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2018ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE(%"class.cl::sycl::accessor"* %0) void acc(sycl::accessor) {} +// CHK-HOST: define dso_local void @_Z3accN2cl4sycl8accessorIiLi1ELNS0_6access4modeE1024ELNS2_6targetE2016ELNS2_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE(%"class.cl::sycl::accessor.3"* %0) +void acc(sycl::accessor) {} + +// CHK-HOST: define dso_local void @_Z3accN2cl4sycl8accessorINS0_3vecIiLi4EEELi1ELNS0_6access4modeE1024ELNS4_6targetE2019ELNS4_11placeholderE0ENS0_6ONEAPI22accessor_property_listIJEEEEE(%"class.cl::sycl::accessor.8"* %0) +void acc(sycl::accessor) {} + // CHK-HOST: define dso_local void @_Z3bufN2cl4sycl6bufferIiLi1ENS0_6detail17aligned_allocatorIcEEvEE(%"class.cl::sycl::buffer"* %0) void buf(sycl::buffer) {} From 6bb297b68b119d3104212ea6fa87395a724cb4ce Mon Sep 17 00:00:00 2001 From: Alexander Batashev Date: Thu, 24 Sep 2020 09:30:41 +0300 Subject: [PATCH 4/4] Only run test on Linux --- sycl/test/abi/user_mangling.cpp | 1 + 1 file changed, 1 insertion(+) diff --git a/sycl/test/abi/user_mangling.cpp b/sycl/test/abi/user_mangling.cpp index d096ee27fa347..e5369d133f9e0 100644 --- a/sycl/test/abi/user_mangling.cpp +++ b/sycl/test/abi/user_mangling.cpp @@ -1,6 +1,7 @@ // clang-format off // RUN: %clangxx -fsycl -c -emit-llvm -S -o - %s | FileCheck %s --check-prefix CHK-HOST // RUN: %clangxx -fsycl -fsycl-device-only -O0 -c -emit-llvm -S -o - %s | FileCheck %s --check-prefix CHK-DEVICE +// REQUIRES: linux #include