diff --git a/clang/test/CodeGenSYCL/accessor_inheritance.cpp b/clang/test/CodeGenSYCL/accessor_inheritance.cpp index 0233b80483b55..f0196d8be0e8e 100644 --- a/clang/test/CodeGenSYCL/accessor_inheritance.cpp +++ b/clang/test/CodeGenSYCL/accessor_inheritance.cpp @@ -23,7 +23,7 @@ int main() { // Check kernel parameters // CHECK: %[[RANGE_TYPE:"struct.*cl::sycl::range"]] // CHECK: %[[ID_TYPE:"struct.*cl::sycl::id"]] -// CHECK: define spir_kernel void @_ZTSZ4mainE6kernel +// CHECK: define {{.*}}spir_kernel void @_ZTSZ4mainE6kernel // CHECK-SAME: i32 [[ARG_A:%[a-zA-Z0-9_]+]], // CHECK-SAME: i32 [[ARG_B:%[a-zA-Z0-9_]+]], // CHECK-SAME: i8 addrspace(1)* [[ACC1_DATA:%[a-zA-Z0-9_]+]], diff --git a/clang/test/CodeGenSYCL/address-space-new.cpp b/clang/test/CodeGenSYCL/address-space-new.cpp index 788d7e0e2f13a..5e332e739b77c 100644 --- a/clang/test/CodeGenSYCL/address-space-new.cpp +++ b/clang/test/CodeGenSYCL/address-space-new.cpp @@ -103,7 +103,7 @@ void test() { // Y yy; baz(yy); - // CHECK: define spir_func void @{{.*}}baz{{.*}} + // CHECK: define {{.*}}spir_func void @{{.*}}baz{{.*}} // CHECK: %[[FIRST:[a-zA-Z0-9]+]] = bitcast %struct.{{.*}}.Y addrspace(4)* %{{.*}} to i8 addrspace(4)* // CHECK: %[[OFFSET:[a-zA-Z0-9]+]].ptr = getelementptr inbounds i8, i8 addrspace(4)* %[[FIRST]], i64 8 // CHECK: %[[SECOND:[a-zA-Z0-9]+]] = bitcast i8 addrspace(4)* %[[OFFSET]].ptr to %struct.{{.*}}.HasX addrspace(4)* diff --git a/clang/test/CodeGenSYCL/address-space-of-returns.cpp b/clang/test/CodeGenSYCL/address-space-of-returns.cpp index 3b56e34bd5bed..a0b477c4e6d51 100644 --- a/clang/test/CodeGenSYCL/address-space-of-returns.cpp +++ b/clang/test/CodeGenSYCL/address-space-of-returns.cpp @@ -26,7 +26,7 @@ A ret_agg() { A a; return a; } -// CHECK: define spir_func void @{{.*}}ret_agg{{.*}}(%struct.{{.*}}.A addrspace(4)* {{.*}} %agg.result) +// CHECK: define {{.*}}spir_func void @{{.*}}ret_agg{{.*}}(%struct.{{.*}}.A addrspace(4)* {{.*}} %agg.result) template __attribute__((sycl_kernel)) void kernel_single_task(const Func &kernelFunc) { diff --git a/clang/test/CodeGenSYCL/address-space-parameter-conversions.cpp b/clang/test/CodeGenSYCL/address-space-parameter-conversions.cpp index cbb645009bb34..cc3ad88b64940 100644 --- a/clang/test/CodeGenSYCL/address-space-parameter-conversions.cpp +++ b/clang/test/CodeGenSYCL/address-space-parameter-conversions.cpp @@ -1,20 +1,20 @@ // RUN: %clang_cc1 -fsycl -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s void bar(int & Data) {} -// CHECK-DAG: define spir_func void @[[RAW_REF:[a-zA-Z0-9_]+]](i32 addrspace(4)* align 4 dereferenceable(4) % +// CHECK-DAG: define {{.*}}spir_func void @[[RAW_REF:[a-zA-Z0-9_]+]](i32 addrspace(4)* align 4 dereferenceable(4) % void bar2(int & Data) {} -// CHECK-DAG: define spir_func void @[[RAW_REF2:[a-zA-Z0-9_]+]](i32 addrspace(4)* align 4 dereferenceable(4) % +// CHECK-DAG: define {{.*}}spir_func void @[[RAW_REF2:[a-zA-Z0-9_]+]](i32 addrspace(4)* align 4 dereferenceable(4) % void bar(__attribute__((opencl_local)) int &Data) {} -// CHECK-DAG: define spir_func void [[LOC_REF:@[a-zA-Z0-9_]+]](i32 addrspace(3)* align 4 dereferenceable(4) % +// CHECK-DAG: define {{.*}}spir_func void [[LOC_REF:@[a-zA-Z0-9_]+]](i32 addrspace(3)* align 4 dereferenceable(4) % void bar3(__attribute__((opencl_global)) int &Data) {} -// CHECK-DAG: define spir_func void @[[GLOB_REF:[a-zA-Z0-9_]+]](i32 addrspace(1)* align 4 dereferenceable(4) % +// CHECK-DAG: define {{.*}}spir_func void @[[GLOB_REF:[a-zA-Z0-9_]+]](i32 addrspace(1)* align 4 dereferenceable(4) % void foo(int * Data) {} -// CHECK-DAG: define spir_func void @[[RAW_PTR:[a-zA-Z0-9_]+]](i32 addrspace(4)* % +// CHECK-DAG: define {{.*}}spir_func void @[[RAW_PTR:[a-zA-Z0-9_]+]](i32 addrspace(4)* % void foo2(int * Data) {} -// CHECK-DAG: define spir_func void @[[RAW_PTR2:[a-zA-Z0-9_]+]](i32 addrspace(4)* % +// CHECK-DAG: define {{.*}}spir_func void @[[RAW_PTR2:[a-zA-Z0-9_]+]](i32 addrspace(4)* % void foo(__attribute__((opencl_local)) int *Data) {} -// CHECK-DAG: define spir_func void [[LOC_PTR:@[a-zA-Z0-9_]+]](i32 addrspace(3)* % +// CHECK-DAG: define {{.*}}spir_func void [[LOC_PTR:@[a-zA-Z0-9_]+]](i32 addrspace(3)* % void foo3(__attribute__((opencl_global)) int *Data) {} -// CHECK-DAG: define spir_func void @[[GLOB_PTR:[a-zA-Z0-9_]+]](i32 addrspace(1)* % +// CHECK-DAG: define {{.*}}spir_func void @[[GLOB_PTR:[a-zA-Z0-9_]+]](i32 addrspace(1)* % template void tmpl(T t){} diff --git a/clang/test/CodeGenSYCL/basic-kernel-wrapper.cpp b/clang/test/CodeGenSYCL/basic-kernel-wrapper.cpp index 750ca8fe0908e..b381dad7b6fe3 100644 --- a/clang/test/CodeGenSYCL/basic-kernel-wrapper.cpp +++ b/clang/test/CodeGenSYCL/basic-kernel-wrapper.cpp @@ -19,7 +19,7 @@ int main() { return 0; } -// CHECK: define spir_kernel void @{{.*}}kernel_function +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_function // CHECK-SAME: i32 addrspace(1)* [[MEM_ARG:%[a-zA-Z0-9_]+]], // CHECK-SAME: %"struct.{{.*}}.cl::sycl::range"* byval{{.*}}align 4 [[ACC_RANGE:%[a-zA-Z0-9_]+_1]], // CHECK-SAME: %"struct.{{.*}}.cl::sycl::range"* byval{{.*}}align 4 [[MEM_RANGE:%[a-zA-Z0-9_]+_2]], diff --git a/clang/test/CodeGenSYCL/device-functions.cpp b/clang/test/CodeGenSYCL/device-functions.cpp index 908c5647c6925..ff5db052a233f 100644 --- a/clang/test/CodeGenSYCL/device-functions.cpp +++ b/clang/test/CodeGenSYCL/device-functions.cpp @@ -21,7 +21,7 @@ int main() { kernel_single_task([]() { foo(); }); return 0; } -// CHECK: define spir_kernel void @_ZTSZ4mainE11fake_kernel() +// CHECK: define {{.*}}spir_kernel void @_ZTSZ4mainE11fake_kernel() // CHECK: define internal spir_func void @"_ZZ4mainENK3$_0clEv"(%"class.{{.*}}.anon" addrspace(4)* {{[^,]*}} %this) -// CHECK: define spir_func void @_Z3foov() +// CHECK: define {{.*}}spir_func void @_Z3foov() // CHECK: define linkonce_odr spir_func i32 @_Z3barIiET_S0_(i32 %arg) diff --git a/clang/test/CodeGenSYCL/device-indirectly-callable-attr.cpp b/clang/test/CodeGenSYCL/device-indirectly-callable-attr.cpp index 5310a067c0e9e..d70ebb277c4e3 100644 --- a/clang/test/CodeGenSYCL/device-indirectly-callable-attr.cpp +++ b/clang/test/CodeGenSYCL/device-indirectly-callable-attr.cpp @@ -7,10 +7,10 @@ void foo() { helper(); } -// CHECK: define spir_func void @{{.*foo.*}}() #[[ATTRS_INDIR_CALL:[0-9]+]] +// CHECK: define {{.*}}spir_func void @{{.*foo.*}}() #[[ATTRS_INDIR_CALL:[0-9]+]] // CHECK: call spir_func void @{{.*helper.*}}() // -// CHECK: define spir_func void @{{.*helper.*}}() #[[ATTRS_NOT_INDIR_CALL:[0-9]+]] +// CHECK: define {{.*}}spir_func void @{{.*helper.*}}() #[[ATTRS_NOT_INDIR_CALL:[0-9]+]] // int bar20(int a) { return a + 20; } @@ -18,7 +18,7 @@ int bar20(int a) { return a + 20; } class A { public: // CHECK-DAG: define linkonce_odr spir_func void @_ZN1A3fooEv{{.*}}#[[ATTRS_INDIR_CALL]] - // CHECK-DAG: define spir_func i32 @_Z5bar20{{.*}}#[[ATTRS_NOT_INDIR_CALL]] + // CHECK-DAG: define {{.*}}spir_func i32 @_Z5bar20{{.*}}#[[ATTRS_NOT_INDIR_CALL]] [[intel::device_indirectly_callable]] void foo() { bar20(10); } // CHECK-DAG: define linkonce_odr spir_func void @_ZN1AC1Ev{{.*}}#[[ATTRS_INDIR_CALL_1:[0-9]+]] diff --git a/clang/test/CodeGenSYCL/emit-kernel-in-virtual-func.cpp b/clang/test/CodeGenSYCL/emit-kernel-in-virtual-func.cpp index 2c4734b509c60..53e88969a3efe 100644 --- a/clang/test/CodeGenSYCL/emit-kernel-in-virtual-func.cpp +++ b/clang/test/CodeGenSYCL/emit-kernel-in-virtual-func.cpp @@ -26,5 +26,4 @@ int main() { } // Ensure that the SPIR-Kernel function is actually emitted. -// CHECK: define spir_kernel void @_ZTSZN7DERIVEDIiE10initializeEvE2FF - +// CHECK: define {{.*}}spir_kernel void @_ZTSZN7DERIVEDIiE10initializeEvE2FF diff --git a/clang/test/CodeGenSYCL/emit-kernel-in-virtual-func2.cpp b/clang/test/CodeGenSYCL/emit-kernel-in-virtual-func2.cpp index 15d11b379fa2d..22cbde55c5839 100644 --- a/clang/test/CodeGenSYCL/emit-kernel-in-virtual-func2.cpp +++ b/clang/test/CodeGenSYCL/emit-kernel-in-virtual-func2.cpp @@ -31,5 +31,4 @@ int main() { } // Ensure that the SPIR-Kernel function is actually emitted. -// CHECK: define spir_kernel void @_ZTSZ2TTIiEvvE2PP - +// CHECK: define {{.*}}spir_kernel void @_ZTSZ2TTIiEvvE2PP diff --git a/clang/test/CodeGenSYCL/esimd-accessor-ptr-md.cpp b/clang/test/CodeGenSYCL/esimd-accessor-ptr-md.cpp index c33526ff5619c..8a9e92b639a18 100644 --- a/clang/test/CodeGenSYCL/esimd-accessor-ptr-md.cpp +++ b/clang/test/CodeGenSYCL/esimd-accessor-ptr-md.cpp @@ -30,7 +30,7 @@ void test(int val) { }); // --- Name - // CHECK-LABEL: define spir_kernel void @"_ZTSZZ4testiENK3$_0clERN2cl4sycl7handlerEE12esimd_kernel"( + // CHECK-LABEL: define {{.*}}spir_kernel void @"_ZTSZZ4testiENK3$_0clERN2cl4sycl7handlerEE12esimd_kernel"( // --- Signature // CHECK: i32 addrspace(1)* "VCArgumentDesc"="buffer_t" "VCArgumentIOKind"="0" "VCArgumentKind"="2" %_arg_, // CHECK: i32 "VCArgumentDesc" "VCArgumentIOKind"="0" "VCArgumentKind"="0" %_arg_1, diff --git a/clang/test/CodeGenSYCL/esimd_metadata1.cpp b/clang/test/CodeGenSYCL/esimd_metadata1.cpp index 39a424c75bf63..5699a68583b3e 100644 --- a/clang/test/CodeGenSYCL/esimd_metadata1.cpp +++ b/clang/test/CodeGenSYCL/esimd_metadata1.cpp @@ -15,7 +15,7 @@ void kernel(const Func &f) __attribute__((sycl_kernel)) { void bar() { kernel([=]() __attribute__((sycl_explicit_simd)){}); - // CHECK: define spir_kernel void @_ZTSZ3barvE8MyKernel() {{.*}} !sycl_explicit_simd ![[EMPTY:[0-9]+]] !intel_reqd_sub_group_size ![[REQD_SIZE:[0-9]+]] + // CHECK: define {{.*}}spir_kernel void @_ZTSZ3barvE8MyKernel() {{.*}} !sycl_explicit_simd ![[EMPTY:[0-9]+]] !intel_reqd_sub_group_size ![[REQD_SIZE:[0-9]+]] } // CHECK: !spirv.Source = !{[[LANG:![0-9]+]]} diff --git a/clang/test/CodeGenSYCL/esimd_metadata2.cpp b/clang/test/CodeGenSYCL/esimd_metadata2.cpp index c14b0af75f3c1..ed88231692327 100644 --- a/clang/test/CodeGenSYCL/esimd_metadata2.cpp +++ b/clang/test/CodeGenSYCL/esimd_metadata2.cpp @@ -11,9 +11,9 @@ __attribute__((sycl_device)) void shared_func() { shared_func_decl(); } __attribute__((sycl_device)) __attribute__((sycl_explicit_simd)) void esimd_func() { shared_func(); } -// CHECK-ESIMD-DAG: define spir_kernel void @{{.*}}kernel_cm() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} {{.*}} !intel_reqd_sub_group_size ![[SGSIZE1:[0-9]+]] {{.*}}{ -// CHECK-ESIMD-DAG: define spir_func void @{{.*}}esimd_funcv() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} { -// CHECK-ESIMD-DAG: define spir_func void @{{.*}}shared_funcv() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} { +// CHECK-ESIMD-DAG: define {{.*}}spir_kernel void @{{.*}}kernel_cm() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} {{.*}} !intel_reqd_sub_group_size ![[SGSIZE1:[0-9]+]] {{.*}}{ +// CHECK-ESIMD-DAG: define {{.*}}spir_func void @{{.*}}esimd_funcv() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} { +// CHECK-ESIMD-DAG: define {{.*}}spir_func void @{{.*}}shared_funcv() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} { // CHECK-ESIMD-DAG: define linkonce_odr spir_func void @_ZN12ESIMDFunctorclEv({{.*}}) #{{[0-9]+}} {{.*}} !sycl_explicit_simd !{{[0-9]+}} { // CHECK-ESIMD-DAG: declare spir_func void @{{.*}}shared_func_declv() #{{[0-9]+}} diff --git a/clang/test/CodeGenSYCL/esimd_metadata3.cpp b/clang/test/CodeGenSYCL/esimd_metadata3.cpp index 02885b1b478db..f34119f4085eb 100644 --- a/clang/test/CodeGenSYCL/esimd_metadata3.cpp +++ b/clang/test/CodeGenSYCL/esimd_metadata3.cpp @@ -3,5 +3,5 @@ __attribute__((sycl_device)) void funcWithSpirvIntrin() {} __attribute__((sycl_device)) __attribute__((sycl_explicit_simd)) void standaloneCmFunc() { funcWithSpirvIntrin(); } -// CHECK-ESIMD-DAG: define spir_func void @{{.*}}funcWithSpirvIntrinv() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} { -// CHECK-ESIMD-DAG: define spir_func void @{{.*}}standaloneCmFuncv() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} { +// CHECK-ESIMD-DAG: define {{.*}}spir_func void @{{.*}}funcWithSpirvIntrinv() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} { +// CHECK-ESIMD-DAG: define {{.*}}spir_func void @{{.*}}standaloneCmFuncv() #{{[0-9]+}} !sycl_explicit_simd !{{[0-9]+}} { diff --git a/clang/test/CodeGenSYCL/image_accessor.cpp b/clang/test/CodeGenSYCL/image_accessor.cpp index 766667b8628dc..817fb7eec7095 100644 --- a/clang/test/CodeGenSYCL/image_accessor.cpp +++ b/clang/test/CodeGenSYCL/image_accessor.cpp @@ -7,27 +7,27 @@ // RUN: FileCheck < %t.ll --enable-var-scope %s --check-prefix=CHECK-3DWO // // CHECK-1DRO: %opencl.image1d_ro_t = type opaque -// CHECK-1DRO: define spir_kernel void @{{.*}}(%opencl.image1d_ro_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) +// CHECK-1DRO: define {{.*}}spir_kernel void @{{.*}}(%opencl.image1d_ro_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) // CHECK-1DRO: call spir_func void @{{.*}}__init{{.*}}(%{{.*}}cl::sycl::accessor{{.*}} %{{[0-9]+}}, %opencl.image1d_ro_t addrspace(1)* %{{[0-9]+}}) // // CHECK-2DRO: %opencl.image2d_ro_t = type opaque -// CHECK-2DRO: define spir_kernel void @{{.*}}(%opencl.image2d_ro_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) +// CHECK-2DRO: define {{.*}}spir_kernel void @{{.*}}(%opencl.image2d_ro_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) // CHECK-2DRO: call spir_func void @{{.*}}__init{{.*}}(%{{.*}}cl::sycl::accessor{{.*}} %{{[0-9]+}}, %opencl.image2d_ro_t addrspace(1)* %{{[0-9]+}}) // // CHECK-3DRO: %opencl.image3d_ro_t = type opaque -// CHECK-3DRO: define spir_kernel void @{{.*}}(%opencl.image3d_ro_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) +// CHECK-3DRO: define {{.*}}spir_kernel void @{{.*}}(%opencl.image3d_ro_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) // CHECK-3DRO: call spir_func void @{{.*}}__init{{.*}}(%{{.*}}cl::sycl::accessor{{.*}} %{{[0-9]+}}, %opencl.image3d_ro_t addrspace(1)* %{{[0-9]+}}) // // CHECK-1DWO: %opencl.image1d_wo_t = type opaque -// CHECK-1DWO: define spir_kernel void @{{.*}}(%opencl.image1d_wo_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) +// CHECK-1DWO: define {{.*}}spir_kernel void @{{.*}}(%opencl.image1d_wo_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) // CHECK-1DWO: call spir_func void @{{.*}}__init{{.*}}(%{{.*}}cl::sycl::accessor{{.*}} %{{[0-9]+}}, %opencl.image1d_wo_t addrspace(1)* %{{[0-9]+}}) // // CHECK-2DWO: %opencl.image2d_wo_t = type opaque -// CHECK-2DWO: define spir_kernel void @{{.*}}(%opencl.image2d_wo_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) +// CHECK-2DWO: define {{.*}}spir_kernel void @{{.*}}(%opencl.image2d_wo_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) // CHECK-2DWO: call spir_func void @{{.*}}__init{{.*}}(%{{.*}}cl::sycl::accessor{{.*}} %{{[0-9]+}}, %opencl.image2d_wo_t addrspace(1)* %{{[0-9]+}}) // // CHECK-3DWO: %opencl.image3d_wo_t = type opaque -// CHECK-3DWO: define spir_kernel void @{{.*}}(%opencl.image3d_wo_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) +// CHECK-3DWO: define {{.*}}spir_kernel void @{{.*}}(%opencl.image3d_wo_t addrspace(1)* [[IMAGE_ARG:%[a-zA-Z0-9_]+]]) // CHECK-3DWO: call spir_func void @{{.*}}__init{{.*}}(%{{.*}}cl::sycl::accessor{{.*}} %{{[0-9]+}}, %opencl.image3d_wo_t addrspace(1)* %{{[0-9]+}}) // // TODO: Add tests for the image_array opencl datatype support. diff --git a/clang/test/CodeGenSYCL/inheritance.cpp b/clang/test/CodeGenSYCL/inheritance.cpp index b5007d1885447..bd5f996c64796 100644 --- a/clang/test/CodeGenSYCL/inheritance.cpp +++ b/clang/test/CodeGenSYCL/inheritance.cpp @@ -40,7 +40,7 @@ int main() { } // Check kernel paramters -// CHECK: define spir_kernel void @{{.*}}derived(%struct.{{.*}}.base* byval(%struct.{{.*}}.base) align 4 %_arg__base, %struct.{{.*}}.__wrapper_class* byval(%struct.{{.*}}.__wrapper_class) align 8 %_arg_e, i32 %_arg_a) +// CHECK: define {{.*}}spir_kernel void @{{.*}}derived(%struct.{{.*}}.base* byval(%struct.{{.*}}.base) align 4 %_arg__base, %struct.{{.*}}.__wrapper_class* byval(%struct.{{.*}}.__wrapper_class) align 8 %_arg_e, i32 %_arg_a) // Check alloca for kernel paramters // CHECK: %[[ARG_A:[a-zA-Z0-9_.]+]] = alloca i32, align 4 diff --git a/clang/test/CodeGenSYCL/intel-fpga-ivdep-array.cpp b/clang/test/CodeGenSYCL/intel-fpga-ivdep-array.cpp index 6ba97282ffecc..9615d3da4a8a0 100644 --- a/clang/test/CodeGenSYCL/intel-fpga-ivdep-array.cpp +++ b/clang/test/CodeGenSYCL/intel-fpga-ivdep-array.cpp @@ -2,7 +2,7 @@ // Array-specific ivdep - annotate the correspondent GEPs only // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_array_no_safelenv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_array_no_safelenv() void ivdep_array_no_safelen() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -18,7 +18,7 @@ void ivdep_array_no_safelen() { } // Array-specific ivdep w/ safelen - annotate the correspondent GEPs only -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_array_with_safelenv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_array_with_safelenv() void ivdep_array_with_safelen() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -35,7 +35,7 @@ void ivdep_array_with_safelen() { // Multiple array-specific ivdeps - annotate the correspondent GEPs // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_multiple_arraysv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_multiple_arraysv() void ivdep_multiple_arrays() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -63,7 +63,7 @@ void ivdep_multiple_arrays() { // Global ivdep with INF safelen & array-specific ivdep with the same safelen // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_array_and_globalv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_array_and_globalv() void ivdep_array_and_global() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -81,7 +81,7 @@ void ivdep_array_and_global() { // Global ivdep with INF safelen & array-specific ivdep with lesser safelen // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_array_and_inf_globalv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_array_and_inf_globalv() void ivdep_array_and_inf_global() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -99,7 +99,7 @@ void ivdep_array_and_inf_global() { // Global ivdep with specified safelen & array-specific ivdep with lesser safelen // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_array_and_greater_globalv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_array_and_greater_globalv() void ivdep_array_and_greater_global() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -117,7 +117,7 @@ void ivdep_array_and_greater_global() { // Global safelen, array-specific safelens // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_mul_arrays_and_globalv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_mul_arrays_and_globalv() void ivdep_mul_arrays_and_global() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -138,7 +138,7 @@ void ivdep_mul_arrays_and_global() { } } -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_ptrv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_ptrv() void ivdep_ptr() { int *ptr; // CHECK: %[[PTR:[0-9a-z]+]] = alloca i32 addrspace(4)* @@ -149,7 +149,7 @@ void ivdep_ptr() { // CHECK: br label %for.cond, !llvm.loop ![[MD_LOOP_PTR:[0-9]+]] } -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_structv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_structv() void ivdep_struct() { struct S { int *ptr; diff --git a/clang/test/CodeGenSYCL/intel-fpga-ivdep-embedded-loops.cpp b/clang/test/CodeGenSYCL/intel-fpga-ivdep-embedded-loops.cpp index 9ce1de58e7738..a1511f0c3886d 100644 --- a/clang/test/CodeGenSYCL/intel-fpga-ivdep-embedded-loops.cpp +++ b/clang/test/CodeGenSYCL/intel-fpga-ivdep-embedded-loops.cpp @@ -2,7 +2,7 @@ // Accesses from the inner loop only, various global safelens for the outer and the inner loops. // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_inner_loop_accessv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_inner_loop_accessv() void ivdep_inner_loop_access() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -19,7 +19,7 @@ void ivdep_inner_loop_access() { // Accesses from both inner and outer loop, same global (INF) safelen for both. // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_embedded_global_safelenv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_embedded_global_safelenv() void ivdep_embedded_global_safelen() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -39,7 +39,7 @@ void ivdep_embedded_global_safelen() { // Accesses from both inner and outer loop, with various safelens per loop. // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_embedded_various_safelensv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_embedded_various_safelensv() void ivdep_embedded_various_safelens() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -61,7 +61,7 @@ void ivdep_embedded_various_safelens() { // Outer loop: array-specific ivdeps for all arrays with various safelens // Inner loop: global ivdep with its own safelen // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_embedded_multiple_arraysv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_embedded_multiple_arraysv() void ivdep_embedded_multiple_arrays() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -90,7 +90,7 @@ void ivdep_embedded_multiple_arrays() { // As the outer loop's ivdep applies to a particular, other array(s) shouldn't be marked // into any index group at the outer loop level // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_embedded_multiple_arrays_globalv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_embedded_multiple_arrays_globalv() void ivdep_embedded_multiple_arrays_global() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -114,7 +114,7 @@ void ivdep_embedded_multiple_arrays_global() { } // Accesses within each dimension of a multi-dimensional (n > 2) loop -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_embedded_multiple_dimensionsv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_embedded_multiple_dimensionsv() void ivdep_embedded_multiple_dimensions() { int a[10]; [[intel::ivdep]] for (int i = 0; i != 10; ++i) { diff --git a/clang/test/CodeGenSYCL/intel-fpga-ivdep-global.cpp b/clang/test/CodeGenSYCL/intel-fpga-ivdep-global.cpp index 84fc0cd0c6372..3ce59a1383510 100644 --- a/clang/test/CodeGenSYCL/intel-fpga-ivdep-global.cpp +++ b/clang/test/CodeGenSYCL/intel-fpga-ivdep-global.cpp @@ -2,7 +2,7 @@ // Global ivdep - annotate all GEPs // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_no_paramv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_no_paramv() void ivdep_no_param() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -20,7 +20,7 @@ void ivdep_no_param() { // Global ivdep - annotate all GEPs // Make sure that ALL of the relevant GEPs for an array are marked into the array's index groups // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_no_param_multiple_gepsv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_no_param_multiple_gepsv() void ivdep_no_param_multiple_geps() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -42,7 +42,7 @@ void ivdep_no_param_multiple_geps() { // Global ivdep w/ safelen specified - annotate all GEPs // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_safelenv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_safelenv() void ivdep_safelen() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; @@ -59,7 +59,7 @@ void ivdep_safelen() { // Global ivdep, albeit conflicting safelens - annotate all GEPs // -// CHECK: define spir_func void @_Z{{[0-9]+}}ivdep_conflicting_safelenv() +// CHECK: define {{.*}}spir_func void @_Z{{[0-9]+}}ivdep_conflicting_safelenv() void ivdep_conflicting_safelen() { // CHECK: %[[ARRAY_A:[0-9a-z]+]] = alloca [10 x i32] int a[10]; diff --git a/clang/test/CodeGenSYCL/intel-fpga-mem-builtin.cpp b/clang/test/CodeGenSYCL/intel-fpga-mem-builtin.cpp index abc75cf10dd85..797d3de3a1199 100644 --- a/clang/test/CodeGenSYCL/intel-fpga-mem-builtin.cpp +++ b/clang/test/CodeGenSYCL/intel-fpga-mem-builtin.cpp @@ -12,7 +12,7 @@ struct State { // CHECK: [[ANN1:@.str[\.]*[0-9]*]] = {{.*}}{params:384}{cache-size:0} // CHECK: [[ANN2:@.str[\.]*[0-9]*]] = {{.*}}{params:384}{cache-size:127} -// CHECK: define spir_func void @{{.*}}(float addrspace(4)* %A, i32 addrspace(4)* %B, [[STRUCT]] addrspace(4)* %C, [[STRUCT]] addrspace(4)*{{.*}}%D) +// CHECK: define {{.*}}spir_func void @{{.*}}(float addrspace(4)* %A, i32 addrspace(4)* %B, [[STRUCT]] addrspace(4)* %C, [[STRUCT]] addrspace(4)*{{.*}}%D) void foo(float *A, int *B, State *C, State &D) { float *x; int *y; diff --git a/clang/test/CodeGenSYCL/intel-fpga-no-global-work-offset.cpp b/clang/test/CodeGenSYCL/intel-fpga-no-global-work-offset.cpp index a2d33c2ac2932..a360337a26bfb 100644 --- a/clang/test/CodeGenSYCL/intel-fpga-no-global-work-offset.cpp +++ b/clang/test/CodeGenSYCL/intel-fpga-no-global-work-offset.cpp @@ -40,10 +40,10 @@ int main() { return 0; } -// CHECK: define spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !no_global_work_offset ![[NUM5:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !no_global_work_offset ![[NUM5]] -// CHECK: define spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} ![[NUM4:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !no_global_work_offset ![[NUM5]] -// CHECK: define spir_kernel void @{{.*}}kernel_name5"() #0 {{.*}} !no_global_work_offset ![[NUM5]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !no_global_work_offset ![[NUM5:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !no_global_work_offset ![[NUM5]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} ![[NUM4:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !no_global_work_offset ![[NUM5]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name5"() #0 {{.*}} !no_global_work_offset ![[NUM5]] // CHECK-NOT: ![[NUM4]] = !{i32 0} // CHECK: ![[NUM5]] = !{} diff --git a/clang/test/CodeGenSYCL/intel-max-global-work-dim.cpp b/clang/test/CodeGenSYCL/intel-max-global-work-dim.cpp index cde0613a8971b..35f0e4d19cf11 100644 --- a/clang/test/CodeGenSYCL/intel-max-global-work-dim.cpp +++ b/clang/test/CodeGenSYCL/intel-max-global-work-dim.cpp @@ -30,8 +30,8 @@ int main() { return 0; } -// CHECK: define spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !max_global_work_dim ![[NUM1:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !max_global_work_dim ![[NUM2:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !max_global_work_dim ![[NUM2]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !max_global_work_dim ![[NUM1:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !max_global_work_dim ![[NUM2:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !max_global_work_dim ![[NUM2]] // CHECK: ![[NUM1]] = !{i32 1} // CHECK: ![[NUM2]] = !{i32 2} diff --git a/clang/test/CodeGenSYCL/intel-max-work-group-size.cpp b/clang/test/CodeGenSYCL/intel-max-work-group-size.cpp index 575177da51106..5700ddfd16c1d 100644 --- a/clang/test/CodeGenSYCL/intel-max-work-group-size.cpp +++ b/clang/test/CodeGenSYCL/intel-max-work-group-size.cpp @@ -45,11 +45,11 @@ int main() { return 0; } -// CHECK: define spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !max_work_group_size ![[NUM1:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !max_work_group_size ![[NUM8:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !max_work_group_size ![[NUM6:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !max_work_group_size ![[NUM2:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name5"() #0 {{.*}} !max_work_group_size ![[NUM4:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !max_work_group_size ![[NUM1:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !max_work_group_size ![[NUM8:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !max_work_group_size ![[NUM6:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !max_work_group_size ![[NUM2:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name5"() #0 {{.*}} !max_work_group_size ![[NUM4:[0-9]+]] // CHECK: ![[NUM1]] = !{i32 1, i32 1, i32 1} // CHECK: ![[NUM8]] = !{i32 8, i32 8, i32 8} // CHECK: ![[NUM6]] = !{i32 6, i32 3, i32 1} diff --git a/clang/test/CodeGenSYCL/intel-restrict.cpp b/clang/test/CodeGenSYCL/intel-restrict.cpp index 12b870ba625f1..9e24c0e85903c 100644 --- a/clang/test/CodeGenSYCL/intel-restrict.cpp +++ b/clang/test/CodeGenSYCL/intel-restrict.cpp @@ -11,7 +11,7 @@ int main() { int *c; kernel( [ a, b, c ]() [[intel::kernel_args_restrict]] { c[0] = a[0] + b[0]; }); - // CHECK: define spir_kernel {{.*}}kernel_restrict(i32 addrspace(1)* noalias %{{.*}}, i32 addrspace(1)* noalias %{{.*}}, i32 addrspace(1)* noalias %{{.*}}) + // CHECK: define {{.*}}spir_kernel {{.*}}kernel_restrict(i32 addrspace(1)* noalias %{{.*}}, i32 addrspace(1)* noalias %{{.*}}, i32 addrspace(1)* noalias %{{.*}}) int *d; int *e; @@ -19,10 +19,10 @@ int main() { kernel( [d, e, f]() { f[0] = d[0] + e[0]; }); - // CHECK: define spir_kernel {{.*}}kernel_norestrict(i32 addrspace(1)* %{{.*}}, i32 addrspace(1)* %{{.*}}, i32 addrspace(1)* %{{.*}}) + // CHECK: define {{.*}}spir_kernel {{.*}}kernel_norestrict(i32 addrspace(1)* %{{.*}}, i32 addrspace(1)* %{{.*}}, i32 addrspace(1)* %{{.*}}) int g = 42; kernel( [ a, b, c, g ]() [[intel::kernel_args_restrict]] { c[0] = a[0] + b[0] + g; }); - // CHECK: define spir_kernel {{.*}}kernel_restrict_other_types(i32 addrspace(1)* noalias %{{.*}}, i32 addrspace(1)* noalias %{{.*}}, i32 addrspace(1)* noalias %{{.*}}, i32 %{{.*}}) + // CHECK: define {{.*}}spir_kernel {{.*}}kernel_restrict_other_types(i32 addrspace(1)* noalias %{{.*}}, i32 addrspace(1)* noalias %{{.*}}, i32 addrspace(1)* noalias %{{.*}}, i32 %{{.*}}) } diff --git a/clang/test/CodeGenSYCL/kernel-name.cpp b/clang/test/CodeGenSYCL/kernel-name.cpp index 167203fd266a7..08c47be78ac01 100644 --- a/clang/test/CodeGenSYCL/kernel-name.cpp +++ b/clang/test/CodeGenSYCL/kernel-name.cpp @@ -13,7 +13,7 @@ __attribute__((sycl_kernel)) void kernel_single_task(const Func &kernelFunc) { kernelFunc(); } -// CHECK: define spir_kernel {{.*}}2cl4sycl6kernel +// CHECK: define {{.*}}spir_kernel {{.*}}2cl4sycl6kernel int main() { kernel_single_task([]() {}); return 0; diff --git a/clang/test/CodeGenSYCL/kernel-param-acc-array.cpp b/clang/test/CodeGenSYCL/kernel-param-acc-array.cpp index a4def6d3df400..2bd2b52945015 100644 --- a/clang/test/CodeGenSYCL/kernel-param-acc-array.cpp +++ b/clang/test/CodeGenSYCL/kernel-param-acc-array.cpp @@ -24,7 +24,7 @@ int main() { } // Check kernel_A parameters -// CHECK: define spir_kernel void @{{.*}}kernel_A +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_A // CHECK-SAME: i32 addrspace(1)* [[MEM_ARG1:%[a-zA-Z0-9_]+]], // CHECK-SAME: %"struct.{{.*}}.cl::sycl::range"* byval{{.*}}align 4 [[ACC_RANGE1:%[a-zA-Z0-9_]+_1]], // CHECK-SAME: %"struct.{{.*}}.cl::sycl::range"* byval{{.*}}align 4 [[MEM_RANGE1:%[a-zA-Z0-9_]+_2]], diff --git a/clang/test/CodeGenSYCL/kernel-param-member-acc-array.cpp b/clang/test/CodeGenSYCL/kernel-param-member-acc-array.cpp index 45d4b658abf83..672e80f0edc38 100644 --- a/clang/test/CodeGenSYCL/kernel-param-member-acc-array.cpp +++ b/clang/test/CodeGenSYCL/kernel-param-member-acc-array.cpp @@ -27,7 +27,7 @@ int main() { } // CHECK kernel_C parameters -// CHECK: define spir_kernel void @{{.*}}kernel_C +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_C // CHECK-SAME: i32 addrspace(1)* [[MEM_ARG1:%[a-zA-Z0-9_]+]], // CHECK-SAME: %"struct.{{.*}}.cl::sycl::range"* byval({{.*}}) align 4 [[ACC_RANGE1:%[a-zA-Z0-9_]+1]], // CHECK-SAME: %"struct.{{.*}}.cl::sycl::range"* byval({{.*}}) align 4 [[MEM_RANGE1:%[a-zA-Z0-9_]+2]], diff --git a/clang/test/CodeGenSYCL/kernel-param-pod-array.cpp b/clang/test/CodeGenSYCL/kernel-param-pod-array.cpp index c9f602261a8f7..4c7ffe0d3c591 100644 --- a/clang/test/CodeGenSYCL/kernel-param-pod-array.cpp +++ b/clang/test/CodeGenSYCL/kernel-param-pod-array.cpp @@ -45,7 +45,7 @@ int main() { } // Check kernel_B parameters -// CHECK: define spir_kernel void @{{.*}}kernel_B +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_B // CHECK-SAME:(%struct.{{.*}}.__wrapper_class* byval(%struct.{{.*}}.__wrapper_class) align 4 %[[ARR_ARG:.*]]) // Check local lambda object alloca @@ -69,7 +69,7 @@ int main() { // CHECK: br i1 %[[ISDONE]], label %{{.*}}, label %[[ARRAYINITBODY]] // Check kernel_C parameters -// CHECK: define spir_kernel void @{{.*}}kernel_C +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_C // CHECK-SAME:(%struct.{{.*}}.__wrapper_class{{.*}}* byval(%struct.{{.*}}.__wrapper_class{{.*}}) align 4 %[[ARR_ARG:.*]]) // Check local lambda object alloca @@ -94,7 +94,7 @@ int main() { // CHECK: br i1 %[[ISDONE]], label %{{.*}}, label %[[ARRAYINITBODY]] // Check kernel_D parameters -// CHECK: define spir_kernel void @{{.*}}kernel_D +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_D // CHECK-SAME:(%struct.{{.*}}.__wrapper_class{{.*}}* byval(%struct.{{.*}}.__wrapper_class{{.*}}) align 4 %[[ARR_ARG:.*]]) // Check local lambda object alloca diff --git a/clang/test/CodeGenSYCL/loop_fuse_device.cpp b/clang/test/CodeGenSYCL/loop_fuse_device.cpp index 2789f05223716..41e281673e1b9 100644 --- a/clang/test/CodeGenSYCL/loop_fuse_device.cpp +++ b/clang/test/CodeGenSYCL/loop_fuse_device.cpp @@ -37,15 +37,15 @@ void bar() { }); } -// CHECK: define spir_kernel void @"{{.*}}kernel_name_1"() {{.*}} !loop_fuse ![[LF5:[0-9]+]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_2"() {{.*}} !loop_fuse ![[LF1:[0-9]+]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_3"() {{.*}} !loop_fuse ![[LF0:[0-9]+]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_4"() {{.*}} !loop_fuse ![[LF1]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_5"() {{.*}} !loop_fuse ![[LF10:[0-9]+]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_6"() +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_1"() {{.*}} !loop_fuse ![[LF5:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_2"() {{.*}} !loop_fuse ![[LF1:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_3"() {{.*}} !loop_fuse ![[LF0:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_4"() {{.*}} !loop_fuse ![[LF1]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_5"() {{.*}} !loop_fuse ![[LF10:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_6"() // CHECK-NOT: !loop_fuse // CHECK-SAME: { -// CHECK: define spir_func void @{{.*}}foo{{.*}} !loop_fuse ![[LF5]] +// CHECK: define {{.*}}spir_func void @{{.*}}foo{{.*}} !loop_fuse ![[LF5]] // CHECK: ![[LF5]] = !{i32 5, i32 0} // CHECK: ![[LF1]] = !{i32 1, i32 0} // CHECK: ![[LF0]] = !{i32 0, i32 0} diff --git a/clang/test/CodeGenSYCL/loop_fuse_ind_device.cpp b/clang/test/CodeGenSYCL/loop_fuse_ind_device.cpp index a415abf74cf44..cab9c90177bc4 100644 --- a/clang/test/CodeGenSYCL/loop_fuse_ind_device.cpp +++ b/clang/test/CodeGenSYCL/loop_fuse_ind_device.cpp @@ -37,15 +37,15 @@ void bar() { }); } -// CHECK: define spir_kernel void @"{{.*}}kernel_name_1"() {{.*}} !loop_fuse ![[LFI5:[0-9]+]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_2"() {{.*}} !loop_fuse ![[LFI1:[0-9]+]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_3"() {{.*}} !loop_fuse ![[LFI0:[0-9]+]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_4"() {{.*}} !loop_fuse ![[LFI1]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_5"() {{.*}} !loop_fuse ![[LFI10:[0-9]+]] -// CHECK: define spir_kernel void @"{{.*}}kernel_name_6"() +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_1"() {{.*}} !loop_fuse ![[LFI5:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_2"() {{.*}} !loop_fuse ![[LFI1:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_3"() {{.*}} !loop_fuse ![[LFI0:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_4"() {{.*}} !loop_fuse ![[LFI1]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_5"() {{.*}} !loop_fuse ![[LFI10:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}kernel_name_6"() // CHECK-NOT: !loop_fuse // CHECK-SAME: { -// CHECK: define spir_func void @{{.*}}foo{{.*}} !loop_fuse ![[LFI5]] +// CHECK: define {{.*}}spir_func void @{{.*}}foo{{.*}} !loop_fuse ![[LFI5]] // CHECK: ![[LFI5]] = !{i32 5, i32 1} // CHECK: ![[LFI1]] = !{i32 1, i32 1} // CHECK: ![[LFI0]] = !{i32 0, i32 1} diff --git a/clang/test/CodeGenSYCL/module-id.cpp b/clang/test/CodeGenSYCL/module-id.cpp index 08ad6178ddbf9..879d6acd0d86b 100644 --- a/clang/test/CodeGenSYCL/module-id.cpp +++ b/clang/test/CodeGenSYCL/module-id.cpp @@ -9,6 +9,6 @@ int main() { kernel_single_task([]() {}); return 0; } -// CHECK: define spir_kernel void @{{.*}}kernel{{.*}}() #[[KERN_ATTR:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel{{.*}}() #[[KERN_ATTR:[0-9]+]] // CHECK: #[[KERN_ATTR]] = { {{.*}}"sycl-module-id"="{{.*}}module-id.cpp"{{.*}} } diff --git a/clang/test/CodeGenSYCL/num-simd-work-items.cpp b/clang/test/CodeGenSYCL/num-simd-work-items.cpp index d83142c29eace..4620ead8b7896 100644 --- a/clang/test/CodeGenSYCL/num-simd-work-items.cpp +++ b/clang/test/CodeGenSYCL/num-simd-work-items.cpp @@ -30,9 +30,9 @@ int main() { return 0; } -// CHECK: define spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !num_simd_work_items ![[NUM1:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !num_simd_work_items ![[NUM42:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !num_simd_work_items ![[NUM2:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !num_simd_work_items ![[NUM1:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !num_simd_work_items ![[NUM42:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !num_simd_work_items ![[NUM2:[0-9]+]] // CHECK: ![[NUM1]] = !{i32 1} // CHECK: ![[NUM42]] = !{i32 42} // CHECK: ![[NUM2]] = !{i32 2} diff --git a/clang/test/CodeGenSYCL/pointers-in-structs.cpp b/clang/test/CodeGenSYCL/pointers-in-structs.cpp index 74d642fb695bf..2f41ad9852cf3 100644 --- a/clang/test/CodeGenSYCL/pointers-in-structs.cpp +++ b/clang/test/CodeGenSYCL/pointers-in-structs.cpp @@ -39,10 +39,10 @@ int main() { // CHECK: %[[WRAPPER_F4_1:[a-zA-Z0-9_.]+]] = type { i32 addrspace(1)* } // CHECK: %[[WRAPPER_F4_2:[a-zA-Z0-9_.]+]] = type { i32 addrspace(1)* } // CHECK: %[[WRAPPER_LAMBDA_PTR:[a-zA-Z0-9_.]+]] = type { float addrspace(1)* } -// CHECK: define spir_kernel void @{{.*}}structs +// CHECK: define {{.*}}spir_kernel void @{{.*}}structs // CHECK-SAME: %[[WRAPPER_F1]]* byval(%[[WRAPPER_F1]]) align 8 %_arg_F1, // CHECK-SAME: %[[WRAPPER_F2]]* byval(%[[WRAPPER_F2]]) align 8 %_arg_F2, // CHECK-SAME: %[[WRAPPER_F]]* byval(%[[WRAPPER_F]]) align 8 %_arg_F, // CHECK-SAME: %[[WRAPPER_F4_1]]* byval(%[[WRAPPER_F4_1]]) align 8 %_arg_F4 // CHECK-SAME: %[[WRAPPER_F4_2]]* byval(%[[WRAPPER_F4_2]]) align 8 %_arg_F41 -// CHECK: define spir_kernel void @{{.*}}lambdas{{.*}}(%[[WRAPPER_LAMBDA_PTR]]* byval(%[[WRAPPER_LAMBDA_PTR]]) align 8 %_arg_) +// CHECK: define {{.*}}spir_kernel void @{{.*}}lambdas{{.*}}(%[[WRAPPER_LAMBDA_PTR]]* byval(%[[WRAPPER_LAMBDA_PTR]]) align 8 %_arg_) diff --git a/clang/test/CodeGenSYCL/reqd-sub-group-size.cpp b/clang/test/CodeGenSYCL/reqd-sub-group-size.cpp index 27bcfeef79ac2..2bd724421a84e 100644 --- a/clang/test/CodeGenSYCL/reqd-sub-group-size.cpp +++ b/clang/test/CodeGenSYCL/reqd-sub-group-size.cpp @@ -42,10 +42,10 @@ int main() { return 0; } -// CHECK: define spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !intel_reqd_sub_group_size ![[SGSIZE16:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !intel_reqd_sub_group_size ![[SGSIZE8:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !intel_reqd_sub_group_size ![[SGSIZE4:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !intel_reqd_sub_group_size ![[SGSIZE2:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !intel_reqd_sub_group_size ![[SGSIZE16:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !intel_reqd_sub_group_size ![[SGSIZE8:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !intel_reqd_sub_group_size ![[SGSIZE4:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !intel_reqd_sub_group_size ![[SGSIZE2:[0-9]+]] // CHECK: ![[SGSIZE16]] = !{i32 16} // CHECK: ![[SGSIZE8]] = !{i32 8} // CHECK: ![[SGSIZE4]] = !{i32 4} diff --git a/clang/test/CodeGenSYCL/reqd-work-group-size.cpp b/clang/test/CodeGenSYCL/reqd-work-group-size.cpp index e0578750af603..230494ee72e51 100644 --- a/clang/test/CodeGenSYCL/reqd-work-group-size.cpp +++ b/clang/test/CodeGenSYCL/reqd-work-group-size.cpp @@ -52,12 +52,12 @@ int main() { return 0; } -// CHECK: define spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE32:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE8:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE88:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE22:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name5"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE44:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name6"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE2:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE32:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE8:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE88:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE22:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name5"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE44:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name6"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE2:[0-9]+]] // CHECK: ![[WGSIZE32]] = !{i32 16, i32 16, i32 32} // CHECK: ![[WGSIZE8]] = !{i32 1, i32 1, i32 8} // CHECK: ![[WGSIZE88]] = !{i32 8, i32 8, i32 8} diff --git a/clang/test/CodeGenSYCL/sampler.cpp b/clang/test/CodeGenSYCL/sampler.cpp index eb0adba66bc09..a5aa3ed837c4e 100644 --- a/clang/test/CodeGenSYCL/sampler.cpp +++ b/clang/test/CodeGenSYCL/sampler.cpp @@ -1,5 +1,5 @@ // RUN: %clang_cc1 -fsycl -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck --enable-var-scope %s -// CHECK: define spir_kernel void @{{[a-zA-Z0-9_]+}}(%opencl.sampler_t addrspace(2)* [[SAMPLER_ARG:%[a-zA-Z0-9_]+]]) +// CHECK: define {{.*}}spir_kernel void @{{[a-zA-Z0-9_]+}}(%opencl.sampler_t addrspace(2)* [[SAMPLER_ARG:%[a-zA-Z0-9_]+]]) // CHECK-NEXT: entry: // CHECK-NEXT: [[SAMPLER_ARG]].addr = alloca %opencl.sampler_t addrspace(2)*, align 8 // CHECK-NEXT: [[ANON:%[0-9]+]] = alloca %"class.{{.*}}.anon", align 8 @@ -12,7 +12,7 @@ // CHECK-NEXT: call spir_func void @{{[a-zA-Z0-9_]+}}(%"class.{{.*}}.cl::sycl::sampler" addrspace(4)* {{[^,]*}} [[GEPCAST]], %opencl.sampler_t addrspace(2)* [[LOAD_SAMPLER_ARG]]) // -// CHECK: define spir_kernel void @{{[a-zA-Z0-9_]+}}(%opencl.sampler_t addrspace(2)* [[SAMPLER_ARG_WRAPPED:%[a-zA-Z0-9_]+]], i32 [[ARG_A:%[a-zA-Z0-9_]+]]) +// CHECK: define {{.*}}spir_kernel void @{{[a-zA-Z0-9_]+}}(%opencl.sampler_t addrspace(2)* [[SAMPLER_ARG_WRAPPED:%[a-zA-Z0-9_]+]], i32 [[ARG_A:%[a-zA-Z0-9_]+]]) // Check alloca // CHECK: [[SAMPLER_ARG_WRAPPED]].addr = alloca %opencl.sampler_t addrspace(2)*, align 8 diff --git a/clang/test/CodeGenSYCL/scheduler-target-fmax-mhz.cpp b/clang/test/CodeGenSYCL/scheduler-target-fmax-mhz.cpp index b5412279a08d0..1fb7fd2513bd2 100644 --- a/clang/test/CodeGenSYCL/scheduler-target-fmax-mhz.cpp +++ b/clang/test/CodeGenSYCL/scheduler-target-fmax-mhz.cpp @@ -17,9 +17,9 @@ int main() { cl::sycl::kernel_single_task( []() { zoo<75>(); }); } -// CHECK: define spir_kernel void @{{.*}}test_kernel1() {{.*}} !scheduler_target_fmax_mhz ![[PARAM1:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}test_kernel2() {{.*}} !scheduler_target_fmax_mhz ![[PARAM2:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}test_kernel3() {{.*}} !scheduler_target_fmax_mhz ![[PARAM3:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}test_kernel1() {{.*}} !scheduler_target_fmax_mhz ![[PARAM1:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}test_kernel2() {{.*}} !scheduler_target_fmax_mhz ![[PARAM2:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}test_kernel3() {{.*}} !scheduler_target_fmax_mhz ![[PARAM3:[0-9]+]] // CHECK: ![[PARAM1]] = !{i32 2} // CHECK: ![[PARAM2]] = !{i32 5} // CHECK: ![[PARAM3]] = !{i32 75} diff --git a/clang/test/CodeGenSYCL/spir-calling-conv.cpp b/clang/test/CodeGenSYCL/spir-calling-conv.cpp index 0da99dc7b167a..2544b4e22f007 100644 --- a/clang/test/CodeGenSYCL/spir-calling-conv.cpp +++ b/clang/test/CodeGenSYCL/spir-calling-conv.cpp @@ -7,7 +7,7 @@ __attribute__((sycl_kernel)) void kernel_single_task(const Func &kernelFunc) { int main() { - // CHECK: define spir_kernel void @_ZTSZ4mainE15kernel_function() + // CHECK: define {{.*}}spir_kernel void @_ZTSZ4mainE15kernel_function() // CHECK: call spir_func void @"_ZZ4mainENK3$_0clEv"(%"class.{{.*}}.anon" addrspace(4)* {{[^,]*}} %2) diff --git a/clang/test/CodeGenSYCL/spir-enum.cpp b/clang/test/CodeGenSYCL/spir-enum.cpp index 81102d213fefa..121270b57642f 100644 --- a/clang/test/CodeGenSYCL/spir-enum.cpp +++ b/clang/test/CodeGenSYCL/spir-enum.cpp @@ -20,7 +20,7 @@ void test(enum_type val) int main() { - // CHECK: define spir_kernel void @_ZTSZ4test9enum_typeE15kernel_function(i32 %_arg_) + // CHECK: define {{.*}}spir_kernel void @_ZTSZ4test9enum_typeE15kernel_function(i32 %_arg_) // CHECK: getelementptr inbounds %"class.{{.*}}.anon", %"class.{{.*}}.anon"* // CHECK: call spir_func void @"_ZZ4test9enum_typeENK3$_0clEv"(%"class.{{.*}}.anon" addrspace(4)* {{[^,]*}} %4) diff --git a/clang/test/CodeGenSYCL/stall_enable.cpp b/clang/test/CodeGenSYCL/stall_enable.cpp index f0e213cc2d219..31acb62f42273 100644 --- a/clang/test/CodeGenSYCL/stall_enable.cpp +++ b/clang/test/CodeGenSYCL/stall_enable.cpp @@ -21,6 +21,6 @@ int main() { return 0; } -// CHECK: define spir_kernel void @"{{.*}}test_kernel1"() #0 {{.*}} !stall_enable ![[NUM5:[0-9]+]] -// CHECK: define spir_kernel void @"{{.*}}test_kernel2"() #0 {{.*}} !stall_enable ![[NUM5]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}test_kernel1"() #0 {{.*}} !stall_enable ![[NUM5:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @"{{.*}}test_kernel2"() #0 {{.*}} !stall_enable ![[NUM5]] // CHECK: ![[NUM5]] = !{i32 1} diff --git a/clang/test/CodeGenSYCL/stream.cpp b/clang/test/CodeGenSYCL/stream.cpp index 476d705054acc..5f2c048438991 100644 --- a/clang/test/CodeGenSYCL/stream.cpp +++ b/clang/test/CodeGenSYCL/stream.cpp @@ -1,7 +1,7 @@ // RUN: %clang_cc1 -fsycl -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o %t.ll // RUN: FileCheck < %t.ll --enable-var-scope %s // -// CHECK: define spir_kernel void @"{{.*}}StreamTester"(%"{{.*}}cl::sycl::stream"* byval(%"{{.*}}cl::sycl::stream") {{.*}}){{.*}} +// CHECK: define {{.*}}spir_kernel void @"{{.*}}StreamTester"(%"{{.*}}cl::sycl::stream"* byval(%"{{.*}}cl::sycl::stream") {{.*}}){{.*}} // CHECK: call spir_func void @{{.*}}__init{{.*}}(%{{.*}}cl::sycl::stream{{.*}} addrspace(4)* {{[^,]*}} %{{[0-9]+}}) // CHECK: call spir_func void @{{.*}}__finalize{{.*}}(%{{.*}}cl::sycl::stream{{.*}} addrspace(4)* {{[^,]*}} %{{[0-9]+}}) // diff --git a/clang/test/CodeGenSYCL/sycl-device-alias.cpp b/clang/test/CodeGenSYCL/sycl-device-alias.cpp index 3a124901b471d..9f8de327c390e 100644 --- a/clang/test/CodeGenSYCL/sycl-device-alias.cpp +++ b/clang/test/CodeGenSYCL/sycl-device-alias.cpp @@ -1,7 +1,7 @@ // RUN: %clang_cc1 -fsycl -fsycl-is-device -triple spir64-unknown-unknown-sycldevice -disable-llvm-passes -emit-llvm %s -o - | FileCheck %s // Test that aliasing does not force an unused entity to be emitted -// CHECK-NOT: define spir_func void @unused_func() +// CHECK-NOT: define {{.*}}spir_func void @unused_func() extern "C" void unused_func() {} // CHECK-NOT: @unused_aliaser extern "C" void unused_aliaser() __attribute__((alias("unused_func"))); @@ -10,29 +10,29 @@ int unused_int = 3; // CHECK-NOT: @alias_unused_int extern int alias_unused_int __attribute__((alias("unused_int"))); -// CHECK-DAG: define spir_func void @used_func() +// CHECK-DAG: define {{.*}}spir_func void @used_func() extern "C" void used_func() {} -// CHECK-DAG: @aliaser = alias void (), void ()* @used_func +// CHECK-DAG: @aliaser = {{.*}}alias void (), void ()* @used_func extern "C" void aliaser() __attribute__((alias("used_func"))); -// CHECK-DAG: define spir_func void @func() +// CHECK-DAG: define {{.*}}spir_func void @func() extern "C" void func() {} -// CHECK-DAG: @used_aliaser = alias void (), void ()* @func +// CHECK-DAG: @used_aliaser = {{.*}}alias void (), void ()* @func extern "C" void used_aliaser() __attribute__((alias("func"))); -// CHECK-DAG: @used_int = addrspace(1) constant i32 5, align 4 +// CHECK-DAG: @used_int = {{.*}}addrspace(1) constant i32 5, align 4 extern "C" const int used_int = 5; -// CHECK-DAG: @alias_used_int = alias i32, i32 addrspace(1)* @used_int +// CHECK-DAG: @alias_used_int = {{.*}}alias i32, i32 addrspace(1)* @used_int extern "C" const int alias_used_int __attribute__((alias("used_int"))); -// CHECK-DAG: @vint = addrspace(1) constant i32 7, align 4 +// CHECK-DAG: @vint = {{.*}}addrspace(1) constant i32 7, align 4 extern "C" const int vint = 7; -// CHECK-DAG: @used_alias_used_int = alias i32, i32 addrspace(1)* @vint +// CHECK-DAG: @used_alias_used_int = {{.*}}alias i32, i32 addrspace(1)* @vint extern "C" const int used_alias_used_int __attribute__((alias("vint"))); -// CHECK-DAG: define spir_func void @{{.*}}bar{{.*}} +// CHECK-DAG: define {{.*}}spir_func void @{{.*}}bar{{.*}} void bar(const int &i) {} -// CHECK-DAG: define spir_func void @{{.*}}foo{{.*}} +// CHECK-DAG: define {{.*}}spir_func void @{{.*}}foo{{.*}} void __attribute__((sycl_device)) foo() { // CHECK-DAG: call spir_func void @{{.*}}bar{{.*}}@used_int bar(used_int); diff --git a/clang/test/CodeGenSYCL/sycl-device-static-init.cpp b/clang/test/CodeGenSYCL/sycl-device-static-init.cpp index 95b220388d894..6b037d632cc56 100644 --- a/clang/test/CodeGenSYCL/sycl-device-static-init.cpp +++ b/clang/test/CodeGenSYCL/sycl-device-static-init.cpp @@ -6,7 +6,7 @@ // CHECK: @_ZN8BaseInitI12TestBaseTypeE3varE = weak_odr addrspace(1) constant i32 9, comdat, align 4 // CHECK-NOT: @_ZN8BaseInitI12TestBaseTypeE15s_regbase_ncsdmE = weak_odr addrspace(1) global %struct._ZTS16RegisterBaseInit.RegisterBaseInit zeroinitializer, comdat, align 1 // CHECK-NOT: @_ZGVN8BaseInitI12TestBaseTypeE15s_regbase_ncsdmE = weak_odr global i64 0, comdat($_ZN8BaseInitI12TestBaseTypeE9s_regbaseE), align 8 -// CHECK: define spir_kernel void @_ZTSZ4mainE11fake_kernel() +// CHECK: define {{.*}}spir_kernel void @_ZTSZ4mainE11fake_kernel() // CHECK: call spir_func void @"_ZZ4mainENK3$_0clEv struct TestBaseType {}; diff --git a/clang/test/CodeGenSYCL/sycl-device.cpp b/clang/test/CodeGenSYCL/sycl-device.cpp index 49f8def476d98..12a363476185b 100644 --- a/clang/test/CodeGenSYCL/sycl-device.cpp +++ b/clang/test/CodeGenSYCL/sycl-device.cpp @@ -9,11 +9,11 @@ int bar20(int a) { return a + 20; } class A { public: // CHECK-DAG: define linkonce_odr spir_func void @_ZN1A3fooEv - // CHECK-DAG: define spir_func i32 @_Z5bar20i + // CHECK-DAG: define {{.*}}spir_func i32 @_Z5bar20i __attribute__((sycl_device)) void foo() { bar20(10); } // CHECK-DAG: define linkonce_odr spir_func void @_ZN1AC1Ev - // CHECK-DAG: define spir_func i32 @_Z5bar10i + // CHECK-DAG: define {{.*}}spir_func i32 @_Z5bar10i __attribute__((sycl_device)) A() { bar10(10); } // CHECK-DAG: define linkonce_odr spir_func void @_ZN1AD1Ev @@ -81,26 +81,26 @@ struct Finalizer1 : Base { void BaseWithAttr() final { int a = 20; } }; -// CHECK-DAG: define spir_func i32 @_Z3fooii +// CHECK-DAG: define {{.*}}spir_func i32 @_Z3fooii __attribute__((sycl_device)) int foo(int a, int b) { return a + bar(b); } -// CHECK-DAG: define spir_func i32 @_Z3bari +// CHECK-DAG: define {{.*}}spir_func i32 @_Z3bari int bar(int b) { return b; } -// CHECK-DAG: define spir_func i32 @_Z3fari +// CHECK-DAG: define {{.*}}spir_func i32 @_Z3fari int far(int b) { return b; } -// CHECK-DAG: define spir_func i32 @_Z3booii +// CHECK-DAG: define {{.*}}spir_func i32 @_Z3booii __attribute__((sycl_device)) int boo(int a, int b) { return a + far(b); } -// CHECK-DAG: define spir_func i32 @_Z3cari +// CHECK-DAG: define {{.*}}spir_func i32 @_Z3cari __attribute__((sycl_device)) int car(int b); int car(int b) { return b; } -// CHECK-DAG: define spir_func i32 @_Z3cazi +// CHECK-DAG: define {{.*}}spir_func i32 @_Z3cazi int caz(int b); __attribute__((sycl_device)) int caz(int b) { return b; } @@ -112,13 +112,13 @@ void taf(T t) {} // CHECK-DAG: define weak_odr spir_func void @_Z3tafIiEvT_ template void taf(int t); -// CHECK-DAG: define spir_func void @_Z3tafIcEvT_ +// CHECK-DAG: define {{.*}}spir_func void @_Z3tafIcEvT_ template<> void taf(char t) {} template void tar(T t) {} -// CHECK-DAG: define spir_func void @_Z3tarIcEvT_ +// CHECK-DAG: define {{.*}}spir_func void @_Z3tarIcEvT_ template<> __attribute__((sycl_device)) void tar(char t) {} diff --git a/clang/test/CodeGenSYCL/sycl-multi-kernel-attr.cpp b/clang/test/CodeGenSYCL/sycl-multi-kernel-attr.cpp index 7fb26c1a8c09d..6f712f38e188a 100644 --- a/clang/test/CodeGenSYCL/sycl-multi-kernel-attr.cpp +++ b/clang/test/CodeGenSYCL/sycl-multi-kernel-attr.cpp @@ -42,10 +42,10 @@ int main() { return 0; } -// CHECK: define spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE:[0-9]+]] !intel_reqd_sub_group_size ![[SGSIZE:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE1:[0-9]+]] !intel_reqd_sub_group_size ![[SGSIZE1:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE2:[0-9]+]] -// CHECK: define spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE3:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name1"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE:[0-9]+]] !intel_reqd_sub_group_size ![[SGSIZE:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name2"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE1:[0-9]+]] !intel_reqd_sub_group_size ![[SGSIZE1:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name3"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE2:[0-9]+]] +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_name4"() #0 {{.*}} !reqd_work_group_size ![[WGSIZE3:[0-9]+]] // CHECK: ![[WGSIZE]] = !{i32 16, i32 16, i32 32} // CHECK: ![[SGSIZE]] = !{i32 4} // CHECK: ![[WGSIZE1]] = !{i32 32, i32 32, i32 64} diff --git a/clang/test/CodeGenSYCL/sycl_kernel-host.cpp b/clang/test/CodeGenSYCL/sycl_kernel-host.cpp index 3414c9deaae8a..e18da6798db6a 100644 --- a/clang/test/CodeGenSYCL/sycl_kernel-host.cpp +++ b/clang/test/CodeGenSYCL/sycl_kernel-host.cpp @@ -4,7 +4,7 @@ // function object passed to the sycl kernel is marked 'alwaysinline' // on the host. -// CHECK: define spir_func void @{{.*}}func{{.*}}() #[[NOSKA:[0-9]+]] { +// CHECK: define {{.*}}spir_func void @{{.*}}func{{.*}}() #[[NOSKA:[0-9]+]] { // CHECK: define internal spir_func void @{{.*}}Kernel{{.*}}Foo{{.*}}({{.*}}) #[[SKA:[0-9]+]] { // CHECK: call spir_func void @{{.*}}KernelImpl{{.*}}({{.*}}, i32 1, double 2.000000e+00) // CHECK: define internal spir_func void @{{.*}}Kernel{{.*}}Bar{{.*}}({{.*}}) #[[SKA]] { diff --git a/clang/test/CodeGenSYCL/union-kernel-param.cpp b/clang/test/CodeGenSYCL/union-kernel-param.cpp index 3cb438802a4f0..45e950b8e48d2 100644 --- a/clang/test/CodeGenSYCL/union-kernel-param.cpp +++ b/clang/test/CodeGenSYCL/union-kernel-param.cpp @@ -28,7 +28,7 @@ int main() { } // CHECK kernel_A parameters -// CHECK: define spir_kernel void @{{.*}}kernel_A(%union.{{.*}}.MyUnion* byval(%union.{{.*}}.MyUnion) align 4 [[MEM_ARG:%[a-zA-Z0-9_]+]]) +// CHECK: define {{.*}}spir_kernel void @{{.*}}kernel_A(%union.{{.*}}.MyUnion* byval(%union.{{.*}}.MyUnion) align 4 [[MEM_ARG:%[a-zA-Z0-9_]+]]) // Check lambda object alloca // CHECK: [[LOCAL_OBJECT:%0]] = alloca %"class.{{.*}}.anon", align 4