[SYCL] Add a test for generated device IR for sycl::vec::as() (#13750)

Author: aelovikov-intel

sycl/test/check_device_code/vector/vector_as.cpp

@@ -0,0 +1,54 @@
+// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --functions "as" --include-generated-funcs --version 4
+// NOTE: "%if preview-breaking-changes-supported" guard has to be temporarily
+// NOTE: removed/disabled to re-generate the checks.
+
+// RUN: %clangxx -O3 -fsycl -fsycl-device-only -fno-discard-value-names -S -emit-llvm -fno-sycl-instrument-device-code  -o - %s | FileCheck %s
+// RUN: %if preview-breaking-changes-supported %{ \
+// RUN: %clangxx -O3 -fsycl -fsycl-device-only -fno-discard-value-names -S -emit-llvm -fno-sycl-instrument-device-code  -o - %s -fpreview-breaking-changes | FileCheck %s --check-prefix=CHECK-PREVIEW \
+// RUN: %}
+
+// Windows/linux have some slight differences in IR generation (function
+// arguments passing and long/long long differences/mangling) that could
+// complicate test updates while not improving test coverage. Limiting to linux
+// should be fine.
+// REQUIRES: linux
+
+#include <sycl/sycl.hpp>
+
+template SYCL_EXTERNAL sycl::vec<int, 4> sycl::vec<float, 4>::as<sycl::vec<int, 4>>() const;
+// CHECK-LABEL: define weak_odr dso_local spir_func void @_ZNK4sycl3_V13vecIfLi4EE2asINS1_IiLi4EEEEET_v(
+// CHECK-SAME: ptr addrspace(4) dead_on_unwind noalias writable sret(%"class.sycl::_V1::vec") align 16 [[AGG_RESULT:%.*]], ptr addrspace(4) noundef align 16 dereferenceable_or_null(16) [[THIS:%.*]]) local_unnamed_addr #[[ATTR0:[0-9]+]] comdat align 2 !srcloc [[META5:![0-9]+]] !sycl_fixed_targets [[META6:![0-9]+]] {
+// CHECK-NEXT:  entry:
+// CHECK-NEXT:    br label [[FOR_COND_I:%.*]]
+// CHECK:       for.cond.i:
+// CHECK-NEXT:    [[I_0_I:%.*]] = phi i64 [ 0, [[ENTRY:%.*]] ], [ [[INC_I:%.*]], [[FOR_BODY_I:%.*]] ]
+// CHECK-NEXT:    [[CMP_I:%.*]] = icmp ult i64 [[I_0_I]], 16
+// CHECK-NEXT:    br i1 [[CMP_I]], label [[FOR_BODY_I]], label [[_ZN4SYCL3_V16DETAIL6MEMCPYEPVPKVM_EXIT:%.*]]
+// CHECK:       for.body.i:
+// CHECK-NEXT:    [[ARRAYIDX_I:%.*]] = getelementptr inbounds i8, ptr addrspace(4) [[THIS]], i64 [[I_0_I]]
+// CHECK-NEXT:    [[TMP0:%.*]] = load i8, ptr addrspace(4) [[ARRAYIDX_I]], align 1, !tbaa [[TBAA7:![0-9]+]]
+// CHECK-NEXT:    [[ARRAYIDX1_I:%.*]] = getelementptr inbounds i8, ptr addrspace(4) [[AGG_RESULT]], i64 [[I_0_I]]
+// CHECK-NEXT:    store i8 [[TMP0]], ptr addrspace(4) [[ARRAYIDX1_I]], align 1, !tbaa [[TBAA7]]
+// CHECK-NEXT:    [[INC_I]] = add nuw nsw i64 [[I_0_I]], 1
+// CHECK-NEXT:    br label [[FOR_COND_I]], !llvm.loop [[LOOP10:![0-9]+]]
+// CHECK:       _ZN4sycl3_V16detail6memcpyEPvPKvm.exit:
+// CHECK-NEXT:    ret void
+//
+//
+// CHECK-PREVIEW-LABEL: define weak_odr dso_local spir_func void @_ZNK4sycl3_V13vecIfLi4EE2asINS1_IiLi4EEEEET_v(
+// CHECK-PREVIEW-SAME: ptr addrspace(4) dead_on_unwind noalias writable sret(%"class.sycl::_V1::vec") align 16 [[AGG_RESULT:%.*]], ptr addrspace(4) noundef align 16 dereferenceable_or_null(16) [[THIS:%.*]]) local_unnamed_addr #[[ATTR0:[0-9]+]] comdat align 2 !srcloc [[META5:![0-9]+]] !sycl_fixed_targets [[META6:![0-9]+]] {
+// CHECK-PREVIEW-NEXT:  entry:
+// CHECK-PREVIEW-NEXT:    br label [[FOR_COND_I:%.*]]
+// CHECK-PREVIEW:       for.cond.i:
+// CHECK-PREVIEW-NEXT:    [[I_0_I:%.*]] = phi i64 [ 0, [[ENTRY:%.*]] ], [ [[INC_I:%.*]], [[FOR_BODY_I:%.*]] ]
+// CHECK-PREVIEW-NEXT:    [[CMP_I:%.*]] = icmp ult i64 [[I_0_I]], 16
+// CHECK-PREVIEW-NEXT:    br i1 [[CMP_I]], label [[FOR_BODY_I]], label [[_ZN4SYCL3_V16DETAIL6MEMCPYEPVPKVM_EXIT:%.*]]
+// CHECK-PREVIEW:       for.body.i:
+// CHECK-PREVIEW-NEXT:    [[ARRAYIDX_I:%.*]] = getelementptr inbounds i8, ptr addrspace(4) [[THIS]], i64 [[I_0_I]]
+// CHECK-PREVIEW-NEXT:    [[TMP0:%.*]] = load i8, ptr addrspace(4) [[ARRAYIDX_I]], align 1, !tbaa [[TBAA7:![0-9]+]]
+// CHECK-PREVIEW-NEXT:    [[ARRAYIDX1_I:%.*]] = getelementptr inbounds i8, ptr addrspace(4) [[AGG_RESULT]], i64 [[I_0_I]]
+// CHECK-PREVIEW-NEXT:    store i8 [[TMP0]], ptr addrspace(4) [[ARRAYIDX1_I]], align 1, !tbaa [[TBAA7]]
+// CHECK-PREVIEW-NEXT:    [[INC_I]] = add nuw nsw i64 [[I_0_I]], 1
+// CHECK-PREVIEW-NEXT:    br label [[FOR_COND_I]], !llvm.loop [[LOOP10:![0-9]+]]
+// CHECK-PREVIEW:       _ZN4sycl3_V16detail6memcpyEPvPKvm.exit:
+// CHECK-PREVIEW-NEXT:    ret void