530 lines · plain
1// REQUIRES: amdgpu-registered-target2// RUN: %clang_cc1 -triple amdgcn-unknown-unknown -emit-llvm -o - %s | FileCheck %s3// RUN: %clang_cc1 -triple r600-unknown-unknown -emit-llvm -o - %s | FileCheck %s4 5typedef __attribute__(( ext_vector_type(2) )) char char2;6typedef __attribute__(( ext_vector_type(3) )) char char3;7typedef __attribute__(( ext_vector_type(4) )) char char4;8 9typedef __attribute__(( ext_vector_type(2) )) short short2;10typedef __attribute__(( ext_vector_type(3) )) short short3;11typedef __attribute__(( ext_vector_type(4) )) short short4;12 13typedef __attribute__(( ext_vector_type(2) )) int int2;14typedef __attribute__(( ext_vector_type(3) )) int int3;15typedef __attribute__(( ext_vector_type(4) )) int int4;16typedef __attribute__(( ext_vector_type(16) )) int int16;17typedef __attribute__(( ext_vector_type(32) )) int int32;18 19// CHECK: %struct.empty_struct = type {}20typedef struct empty_struct21{22} empty_struct;23 24// CHECK-NOT: %struct.single_element_struct_arg25typedef struct single_element_struct_arg26{27 int i;28} single_element_struct_arg_t;29 30// CHECK-NOT: %struct.nested_single_element_struct_arg31typedef struct nested_single_element_struct_arg32{33 single_element_struct_arg_t i;34} nested_single_element_struct_arg_t;35 36// CHECK: %struct.struct_arg = type { i32, float, i32 }37typedef struct struct_arg38{39 int i1;40 float f;41 int i2;42} struct_arg_t;43 44// CHECK: %struct.struct_padding_arg = type { i8, i64 }45typedef struct struct_padding_arg46{47 char i1;48 long f;49} struct_padding_arg;50 51// CHECK: %struct.struct_of_arrays_arg = type { [2 x i32], float, [4 x i32], [3 x float], i32 }52typedef struct struct_of_arrays_arg53{54 int i1[2];55 float f1;56 int i2[4];57 float f2[3];58 int i3;59} struct_of_arrays_arg_t;60 61// CHECK: %struct.struct_of_structs_arg = type { i32, float, %struct.struct_arg, i32 }62typedef struct struct_of_structs_arg63{64 int i1;65 float f1;66 struct_arg_t s1;67 int i2;68} struct_of_structs_arg_t;69 70typedef union71{72 int b1;73 float b2;74} transparent_u __attribute__((__transparent_union__));75 76// CHECK: %struct.single_array_element_struct_arg = type { [4 x i32] }77typedef struct single_array_element_struct_arg78{79 int i[4];80} single_array_element_struct_arg_t;81 82// CHECK: %struct.single_struct_element_struct_arg = type { %struct.inner }83// CHECK: %struct.inner = type { i32, i64 }84typedef struct single_struct_element_struct_arg85{86 struct inner {87 int a;88 long b;89 } s;90} single_struct_element_struct_arg_t;91 92// CHECK: %struct.different_size_type_pair93typedef struct different_size_type_pair {94 long l;95 int i;96} different_size_type_pair;97 98// CHECK: %struct.flexible_array = type { i32, [0 x i32] }99typedef struct flexible_array100{101 int i;102 int flexible[];103} flexible_array;104 105// CHECK: %struct.struct_arr16 = type { [16 x i32] }106typedef struct struct_arr16107{108 int arr[16];109} struct_arr16;110 111// CHECK: %struct.struct_arr32 = type { [32 x i32] }112typedef struct struct_arr32113{114 int arr[32];115} struct_arr32;116 117// CHECK: %struct.struct_arr33 = type { [33 x i32] }118typedef struct struct_arr33119{120 int arr[33];121} struct_arr33;122 123// CHECK: %struct.struct_char_arr32 = type { [32 x i8] }124typedef struct struct_char_arr32125{126 char arr[32];127} struct_char_arr32;128 129// CHECK-NOT: %struct.struct_char_x8130typedef struct struct_char_x8 {131 char x, y, z, w;132 char a, b, c, d;133} struct_char_x8;134 135// CHECK-NOT: %struct.struct_char_x4136typedef struct struct_char_x4 {137 char x, y, z, w;138} struct_char_x4;139 140// CHECK-NOT: %struct.struct_char_x3141typedef struct struct_char_x3 {142 char x, y, z;143} struct_char_x3;144 145// CHECK-NOT: %struct.struct_char_x2146typedef struct struct_char_x2 {147 char x, y;148} struct_char_x2;149 150// CHECK-NOT: %struct.struct_char_x1151typedef struct struct_char_x1 {152 char x;153} struct_char_x1;154 155// 4 registers from fields, 5 if padding included.156// CHECK: %struct.nested = type { i8, i64 }157// CHECK: %struct.num_regs_nested_struct = type { i32, %struct.nested }158typedef struct num_regs_nested_struct {159 int x;160 struct nested {161 char z;162 long y;163 } inner;164} num_regs_nested_struct;165 166// CHECK: %struct.double_nested = type { %struct.inner_inner }167// CHECK: %struct.inner_inner = type { i8, i32, i8 }168// CHECK: %struct.double_nested_struct = type { i32, %struct.double_nested, i16 }169typedef struct double_nested_struct {170 int x;171 struct double_nested {172 struct inner_inner {173 char y;174 int q;175 char z;176 } inner_inner;177 } inner;178 179 short w;180} double_nested_struct;181 182// This is a large struct, but uses fewer registers than the limit.183// CHECK: %struct.large_struct_padding = type { i8, i32, i8, i32, i8, i8, i16, i16, [3 x i8], i64, i32, i8, i32, i16, i8 }184typedef struct large_struct_padding {185 char e0;186 int e1;187 char e2;188 int e3;189 char e4;190 char e5;191 short e6;192 short e7;193 char e8[3];194 long e9;195 int e10;196 char e11;197 int e12;198 short e13;199 char e14;200} large_struct_padding;201 202// The number of registers computed should be 6, not 8.203typedef struct int3_pair {204 int3 dx;205 int3 dy;206} int3_pair;207 208// CHECK: %struct.struct_4regs = type { i32, i32, i32, i32 }209typedef struct struct_4regs210{211 int x;212 int y;213 int z;214 int w;215} struct_4regs;216 217// CHECK: void @kernel_empty_struct_arg(ptr addrspace(4) noundef readnone byref(%struct.empty_struct) align 1 captures(none) {{%.+}})218// CHECK: void @__clang_ocl_kern_imp_kernel_empty_struct_arg()219__kernel void kernel_empty_struct_arg(empty_struct s) { }220 221// CHECK: void @kernel_single_element_struct_arg(i32 %arg1.coerce)222__kernel void kernel_single_element_struct_arg(single_element_struct_arg_t arg1) { }223 224// CHECK: void @kernel_nested_single_element_struct_arg(i32 %arg1.coerce)225__kernel void kernel_nested_single_element_struct_arg(nested_single_element_struct_arg_t arg1) { }226 227// CHECK: void @kernel_struct_arg(ptr addrspace(4) noundef readonly byref(%struct.struct_arg) align 4 captures(none) {{%.+}})228// CHECK: void @__clang_ocl_kern_imp_kernel_struct_arg(i32 %arg1.coerce0, float %arg1.coerce1, i32 %arg1.coerce2)229__kernel void kernel_struct_arg(struct_arg_t arg1) { }230 231// CHECK: void @kernel_struct_padding_arg(ptr addrspace(4) noundef readonly byref(%struct.struct_padding_arg) align 8 captures(none) {{%.+}})232// CHECK: void @__clang_ocl_kern_imp_kernel_struct_padding_arg(i8 %arg1.coerce0, i64 %arg1.coerce1)233__kernel void kernel_struct_padding_arg(struct_padding_arg arg1) { }234 235// CHECK: void @kernel_test_struct_of_arrays_arg(ptr addrspace(4) noundef readonly byref(%struct.struct_of_arrays_arg) align 4 captures(none) {{%.+}})236// CHECK: void @__clang_ocl_kern_imp_kernel_test_struct_of_arrays_arg([2 x i32] %arg1.coerce0, float %arg1.coerce1, [4 x i32] %arg1.coerce2, [3 x float] %arg1.coerce3, i32 %arg1.coerce4)237__kernel void kernel_test_struct_of_arrays_arg(struct_of_arrays_arg_t arg1) { }238 239// CHECK: void @kernel_struct_of_structs_arg(ptr addrspace(4) noundef readonly byref(%struct.struct_of_structs_arg) align 4 captures(none) {{%.+}})240// CHECK: void @__clang_ocl_kern_imp_kernel_struct_of_structs_arg(i32 %arg1.coerce0, float %arg1.coerce1, %struct.struct_arg %arg1.coerce2, i32 %arg1.coerce3)241__kernel void kernel_struct_of_structs_arg(struct_of_structs_arg_t arg1) { }242 243// CHECK: void @test_kernel_transparent_union_arg(i32 %u.coerce)244__kernel void test_kernel_transparent_union_arg(transparent_u u) { }245 246// CHECK: void @kernel_single_array_element_struct_arg(ptr addrspace(4) noundef readonly byref(%struct.single_array_element_struct_arg) align 4 captures(none) {{%.+}})247// CHECK: void @__clang_ocl_kern_imp_kernel_single_array_element_struct_arg([4 x i32] %arg1.coerce)248__kernel void kernel_single_array_element_struct_arg(single_array_element_struct_arg_t arg1) { }249 250// CHECK: void @kernel_single_struct_element_struct_arg(ptr addrspace(4) noundef readonly byref(%struct.single_struct_element_struct_arg) align 8 captures(none) {{%.+}})251// CHECK: void @__clang_ocl_kern_imp_kernel_single_struct_element_struct_arg(%struct.inner %arg1.coerce)252__kernel void kernel_single_struct_element_struct_arg(single_struct_element_struct_arg_t arg1) { }253 254// CHECK: void @kernel_different_size_type_pair_arg(ptr addrspace(4) noundef readonly byref(%struct.different_size_type_pair) align 8 captures(none) {{%.+}})255// CHECK: void @__clang_ocl_kern_imp_kernel_different_size_type_pair_arg(i64 %arg1.coerce0, i32 %arg1.coerce1)256__kernel void kernel_different_size_type_pair_arg(different_size_type_pair arg1) { }257 258// CHECK: define{{.*}} void @func_f32_arg(float noundef %arg)259void func_f32_arg(float arg) { }260 261// CHECK: define{{.*}} void @func_v2i16_arg(<2 x i16> noundef %arg)262void func_v2i16_arg(short2 arg) { }263 264// CHECK: define{{.*}} void @func_v3i32_arg(<3 x i32> noundef %arg)265void func_v3i32_arg(int3 arg) { }266 267// CHECK: define{{.*}} void @func_v4i32_arg(<4 x i32> noundef %arg)268void func_v4i32_arg(int4 arg) { }269 270// CHECK: define{{.*}} void @func_v16i32_arg(<16 x i32> noundef %arg)271void func_v16i32_arg(int16 arg) { }272 273// CHECK: define{{.*}} void @func_v32i32_arg(<32 x i32> noundef %arg)274void func_v32i32_arg(int32 arg) { }275 276// CHECK: define{{.*}} void @func_empty_struct_arg()277void func_empty_struct_arg(empty_struct empty) { }278 279// CHECK: void @func_single_element_struct_arg(i32 %arg1.coerce)280void func_single_element_struct_arg(single_element_struct_arg_t arg1) { }281 282// CHECK: void @func_nested_single_element_struct_arg(i32 %arg1.coerce)283void func_nested_single_element_struct_arg(nested_single_element_struct_arg_t arg1) { }284 285// CHECK: void @func_struct_arg(i32 %arg1.coerce0, float %arg1.coerce1, i32 %arg1.coerce2)286void func_struct_arg(struct_arg_t arg1) { }287 288// CHECK: void @func_struct_padding_arg(i8 %arg1.coerce0, i64 %arg1.coerce1)289void func_struct_padding_arg(struct_padding_arg arg1) { }290 291// CHECK: define{{.*}} void @func_struct_char_x8([2 x i32] %arg.coerce)292void func_struct_char_x8(struct_char_x8 arg) { }293 294// CHECK: define{{.*}} void @func_struct_char_x4(i32 %arg.coerce)295void func_struct_char_x4(struct_char_x4 arg) { }296 297// CHECK: define{{.*}} void @func_struct_char_x3(i32 %arg.coerce)298void func_struct_char_x3(struct_char_x3 arg) { }299 300// CHECK: define{{.*}} void @func_struct_char_x2(i16 %arg.coerce)301void func_struct_char_x2(struct_char_x2 arg) { }302 303// CHECK: define{{.*}} void @func_struct_char_x1(i8 %arg.coerce)304void func_struct_char_x1(struct_char_x1 arg) { }305 306// CHECK: void @func_transparent_union_arg(i32 %u.coerce)307void func_transparent_union_arg(transparent_u u) { }308 309// CHECK: void @func_single_array_element_struct_arg([4 x i32] %arg1.coerce)310void func_single_array_element_struct_arg(single_array_element_struct_arg_t arg1) { }311 312// CHECK: void @func_single_struct_element_struct_arg(%struct.inner %arg1.coerce)313void func_single_struct_element_struct_arg(single_struct_element_struct_arg_t arg1) { }314 315// CHECK: void @func_different_size_type_pair_arg(i64 %arg1.coerce0, i32 %arg1.coerce1)316void func_different_size_type_pair_arg(different_size_type_pair arg1) { }317 318// CHECK: void @func_flexible_array_arg(ptr addrspace(5) noundef readnone byval(%struct.flexible_array) align 4 captures(none) %arg)319void func_flexible_array_arg(flexible_array arg) { }320 321// CHECK: define{{.*}} float @func_f32_ret()322float func_f32_ret()323{324 return 0.0f;325}326 327// CHECK: define{{.*}} void @func_empty_struct_ret()328empty_struct func_empty_struct_ret()329{330 empty_struct s = {};331 return s;332}333 334// CHECK: define{{.*}} i32 @single_element_struct_ret()335// CHECK: ret i32 0336single_element_struct_arg_t single_element_struct_ret()337{338 single_element_struct_arg_t s = { 0 };339 return s;340}341 342// CHECK: define{{.*}} i32 @nested_single_element_struct_ret()343// CHECK: ret i32 0344nested_single_element_struct_arg_t nested_single_element_struct_ret()345{346 nested_single_element_struct_arg_t s = { 0 };347 return s;348}349 350// CHECK: define{{.*}} %struct.struct_arg @func_struct_ret()351// CHECK: ret %struct.struct_arg zeroinitializer352struct_arg_t func_struct_ret()353{354 struct_arg_t s = { 0 };355 return s;356}357 358// CHECK: define{{.*}} %struct.struct_padding_arg @func_struct_padding_ret()359// CHECK: ret %struct.struct_padding_arg zeroinitializer360struct_padding_arg func_struct_padding_ret()361{362 struct_padding_arg s = { 0 };363 return s;364}365 366// CHECK: define{{.*}} [2 x i32] @func_struct_char_x8_ret()367// CHECK: ret [2 x i32] zeroinitializer368struct_char_x8 func_struct_char_x8_ret()369{370 struct_char_x8 s = { 0 };371 return s;372}373 374// CHECK: define{{.*}} i32 @func_struct_char_x4_ret()375// CHECK: ret i32 0376struct_char_x4 func_struct_char_x4_ret()377{378 struct_char_x4 s = { 0 };379 return s;380}381 382// CHECK: define{{.*}} i32 @func_struct_char_x3_ret()383// CHECK: ret i32 0384struct_char_x3 func_struct_char_x3_ret()385{386 struct_char_x3 s = { 0 };387 return s;388}389 390// CHECK: define{{.*}} i16 @func_struct_char_x2_ret()391struct_char_x2 func_struct_char_x2_ret()392{393 struct_char_x2 s = { 0 };394 return s;395}396 397// CHECK: define{{.*}} i8 @func_struct_char_x1_ret()398// CHECK: ret i8 0399struct_char_x1 func_struct_char_x1_ret()400{401 struct_char_x1 s = { 0 };402 return s;403}404 405// CHECK: define{{.*}} %struct.struct_arr16 @func_ret_struct_arr16()406// CHECK: ret %struct.struct_arr16 zeroinitializer407struct_arr16 func_ret_struct_arr16()408{409 struct_arr16 s = { 0 };410 return s;411}412 413// CHECK: define{{.*}} void @func_ret_struct_arr32(ptr addrspace(5) dead_on_unwind noalias writable writeonly sret(%struct.struct_arr32) align 4 captures(none) initializes((0, 128)) %agg.result)414struct_arr32 func_ret_struct_arr32()415{416 struct_arr32 s = { 0 };417 return s;418}419 420// CHECK: define{{.*}} void @func_ret_struct_arr33(ptr addrspace(5) dead_on_unwind noalias writable writeonly sret(%struct.struct_arr33) align 4 captures(none) initializes((0, 132)) %agg.result)421struct_arr33 func_ret_struct_arr33()422{423 struct_arr33 s = { 0 };424 return s;425}426 427// CHECK: define{{.*}} %struct.struct_char_arr32 @func_ret_struct_char_arr32()428struct_char_arr32 func_ret_struct_char_arr32()429{430 struct_char_arr32 s = { 0 };431 return s;432}433 434// CHECK: define{{.*}} i32 @func_transparent_union_ret() local_unnamed_addr #[[ATTR1:[0-9]+]] {435// CHECK: ret i32 0436transparent_u func_transparent_union_ret()437{438 transparent_u u = { 0 };439 return u;440}441 442// CHECK: define{{.*}} %struct.different_size_type_pair @func_different_size_type_pair_ret()443different_size_type_pair func_different_size_type_pair_ret()444{445 different_size_type_pair s = { 0 };446 return s;447}448 449// CHECK: define{{.*}} void @func_flexible_array_ret(ptr addrspace(5) dead_on_unwind noalias writable writeonly sret(%struct.flexible_array) align 4 captures(none) initializes((0, 4)) %agg.result)450flexible_array func_flexible_array_ret()451{452 flexible_array s = { 0 };453 return s;454}455 456// CHECK: define{{.*}} void @func_reg_state_lo(<4 x i32> noundef %arg0, <4 x i32> noundef %arg1, <4 x i32> noundef %arg2, i32 noundef %arg3, i32 %s.coerce0, float %s.coerce1, i32 %s.coerce2)457void func_reg_state_lo(int4 arg0, int4 arg1, int4 arg2, int arg3, struct_arg_t s) { }458 459// CHECK: define{{.*}} void @func_reg_state_hi(<4 x i32> noundef %arg0, <4 x i32> noundef %arg1, <4 x i32> noundef %arg2, i32 noundef %arg3, i32 noundef %arg4, ptr addrspace(5) noundef readnone byref(%struct.struct_arg) align 4 captures(none) %{{.*}})460void func_reg_state_hi(int4 arg0, int4 arg1, int4 arg2, int arg3, int arg4, struct_arg_t s) { }461 462// XXX - Why don't the inner structs flatten?463// CHECK: define{{.*}} void @func_reg_state_num_regs_nested_struct(<4 x i32> noundef %arg0, i32 noundef %arg1, i32 %arg2.coerce0, %struct.nested %arg2.coerce1, i32 %arg3.coerce0, %struct.nested %arg3.coerce1, ptr addrspace(5) noundef readnone byref(%struct.num_regs_nested_struct) align 8 captures(none) %{{.*}})464void func_reg_state_num_regs_nested_struct(int4 arg0, int arg1, num_regs_nested_struct arg2, num_regs_nested_struct arg3, num_regs_nested_struct arg4) { }465 466// CHECK: define{{.*}} void @func_double_nested_struct_arg(<4 x i32> noundef %arg0, i32 noundef %arg1, i32 %arg2.coerce0, %struct.double_nested %arg2.coerce1, i16 %arg2.coerce2)467void func_double_nested_struct_arg(int4 arg0, int arg1, double_nested_struct arg2) { }468 469// CHECK: define{{.*}} %struct.double_nested_struct @func_double_nested_struct_ret(<4 x i32> noundef %arg0, i32 noundef %arg1)470double_nested_struct func_double_nested_struct_ret(int4 arg0, int arg1) {471 double_nested_struct s = { 0 };472 return s;473}474 475// CHECK: define{{.*}} void @func_large_struct_padding_arg_direct(i8 %arg.coerce0, i32 %arg.coerce1, i8 %arg.coerce2, i32 %arg.coerce3, i8 %arg.coerce4, i8 %arg.coerce5, i16 %arg.coerce6, i16 %arg.coerce7, [3 x i8] %arg.coerce8, i64 %arg.coerce9, i32 %arg.coerce10, i8 %arg.coerce11, i32 %arg.coerce12, i16 %arg.coerce13, i8 %arg.coerce14)476void func_large_struct_padding_arg_direct(large_struct_padding arg) { }477 478// CHECK: define{{.*}} void @func_large_struct_padding_arg_store(ptr addrspace(1) noundef writeonly captures(none) initializes((0, 56)) %out, ptr addrspace(5) noundef readonly byref(%struct.large_struct_padding) align 8 captures(none) %{{.*}})479void func_large_struct_padding_arg_store(global large_struct_padding* out, large_struct_padding arg) {480 *out = arg;481}482 483// CHECK: define{{.*}} void @v3i32_reg_count(<3 x i32> noundef %arg1, <3 x i32> noundef %arg2, <3 x i32> noundef %arg3, <3 x i32> noundef %arg4, i32 %arg5.coerce0, float %arg5.coerce1, i32 %arg5.coerce2)484void v3i32_reg_count(int3 arg1, int3 arg2, int3 arg3, int3 arg4, struct_arg_t arg5) { }485 486// Function signature from blender, nothing should be passed byval. The v3i32487// should not count as 4 passed registers.488// CHECK: define{{.*}} void @v3i32_pair_reg_count(ptr addrspace(5) noundef readnone captures(none) %arg0, <3 x i32> %arg1.coerce0, <3 x i32> %arg1.coerce1, <3 x i32> noundef %arg2, <3 x i32> %arg3.coerce0, <3 x i32> %arg3.coerce1, <3 x i32> noundef %arg4, float noundef %arg5)489void v3i32_pair_reg_count(int3_pair *arg0, int3_pair arg1, int3 arg2, int3_pair arg3, int3 arg4, float arg5) { }490 491// Each short4 should fit pack into 2 registers.492// CHECK: define{{.*}} void @v4i16_reg_count(<4 x i16> noundef %arg0, <4 x i16> noundef %arg1, <4 x i16> noundef %arg2, <4 x i16> noundef %arg3, <4 x i16> noundef %arg4, <4 x i16> noundef %arg5, i32 %arg6.coerce0, i32 %arg6.coerce1, i32 %arg6.coerce2, i32 %arg6.coerce3)493void v4i16_reg_count(short4 arg0, short4 arg1, short4 arg2, short4 arg3,494 short4 arg4, short4 arg5, struct_4regs arg6) { }495 496// CHECK: define{{.*}} void @v4i16_pair_reg_count_over(<4 x i16> noundef %arg0, <4 x i16> noundef %arg1, <4 x i16> noundef %arg2, <4 x i16> noundef %arg3, <4 x i16> noundef %arg4, <4 x i16> noundef %arg5, <4 x i16> noundef %arg6, ptr addrspace(5) noundef readnone byref(%struct.struct_4regs) align 4 captures(none) %{{.*}})497void v4i16_pair_reg_count_over(short4 arg0, short4 arg1, short4 arg2, short4 arg3,498 short4 arg4, short4 arg5, short4 arg6, struct_4regs arg7) { }499 500// CHECK: define{{.*}} void @v3i16_reg_count(<3 x i16> noundef %arg0, <3 x i16> noundef %arg1, <3 x i16> noundef %arg2, <3 x i16> noundef %arg3, <3 x i16> noundef %arg4, <3 x i16> noundef %arg5, i32 %arg6.coerce0, i32 %arg6.coerce1, i32 %arg6.coerce2, i32 %arg6.coerce3)501void v3i16_reg_count(short3 arg0, short3 arg1, short3 arg2, short3 arg3,502 short3 arg4, short3 arg5, struct_4regs arg6) { }503 504// CHECK: define{{.*}} void @v3i16_reg_count_over(<3 x i16> noundef %arg0, <3 x i16> noundef %arg1, <3 x i16> noundef %arg2, <3 x i16> noundef %arg3, <3 x i16> noundef %arg4, <3 x i16> noundef %arg5, <3 x i16> noundef %arg6, ptr addrspace(5) noundef readnone byref(%struct.struct_4regs) align 4 captures(none) %{{.*}})505void v3i16_reg_count_over(short3 arg0, short3 arg1, short3 arg2, short3 arg3,506 short3 arg4, short3 arg5, short3 arg6, struct_4regs arg7) { }507 508// CHECK: define{{.*}} void @v2i16_reg_count(<2 x i16> noundef %arg0, <2 x i16> noundef %arg1, <2 x i16> noundef %arg2, <2 x i16> noundef %arg3, <2 x i16> noundef %arg4, <2 x i16> noundef %arg5, <2 x i16> noundef %arg6, <2 x i16> noundef %arg7, <2 x i16> noundef %arg8, <2 x i16> noundef %arg9, <2 x i16> noundef %arg10, <2 x i16> noundef %arg11, i32 %arg13.coerce0, i32 %arg13.coerce1, i32 %arg13.coerce2, i32 %arg13.coerce3)509void v2i16_reg_count(short2 arg0, short2 arg1, short2 arg2, short2 arg3,510 short2 arg4, short2 arg5, short2 arg6, short2 arg7,511 short2 arg8, short2 arg9, short2 arg10, short2 arg11,512 struct_4regs arg13) { }513 514// CHECK: define{{.*}} void @v2i16_reg_count_over(<2 x i16> noundef %arg0, <2 x i16> noundef %arg1, <2 x i16> noundef %arg2, <2 x i16> noundef %arg3, <2 x i16> noundef %arg4, <2 x i16> noundef %arg5, <2 x i16> noundef %arg6, <2 x i16> noundef %arg7, <2 x i16> noundef %arg8, <2 x i16> noundef %arg9, <2 x i16> noundef %arg10, <2 x i16> noundef %arg11, <2 x i16> noundef %arg12, ptr addrspace(5) noundef readnone byref(%struct.struct_4regs) align 4 captures(none) %{{.*}})515void v2i16_reg_count_over(short2 arg0, short2 arg1, short2 arg2, short2 arg3,516 short2 arg4, short2 arg5, short2 arg6, short2 arg7,517 short2 arg8, short2 arg9, short2 arg10, short2 arg11,518 short2 arg12, struct_4regs arg13) { }519 520// CHECK: define{{.*}} void @v2i8_reg_count(<2 x i8> noundef %arg0, <2 x i8> noundef %arg1, <2 x i8> noundef %arg2, <2 x i8> noundef %arg3, <2 x i8> noundef %arg4, <2 x i8> noundef %arg5, i32 %arg6.coerce0, i32 %arg6.coerce1, i32 %arg6.coerce2, i32 %arg6.coerce3)521void v2i8_reg_count(char2 arg0, char2 arg1, char2 arg2, char2 arg3,522 char2 arg4, char2 arg5, struct_4regs arg6) { }523 524// CHECK: define{{.*}} void @v2i8_reg_count_over(<2 x i8> noundef %arg0, <2 x i8> noundef %arg1, <2 x i8> noundef %arg2, <2 x i8> noundef %arg3, <2 x i8> noundef %arg4, <2 x i8> noundef %arg5, i32 noundef %arg6, ptr addrspace(5) noundef readnone byref(%struct.struct_4regs) align 4 captures(none) %{{.*}})525void v2i8_reg_count_over(char2 arg0, char2 arg1, char2 arg2, char2 arg3,526 char2 arg4, char2 arg5, int arg6, struct_4regs arg7) { }527 528// CHECK: define{{.*}} void @num_regs_left_64bit_aggregate(<4 x i32> noundef %arg0, <4 x i32> noundef %arg1, <4 x i32> noundef %arg2, <3 x i32> noundef %arg3, [2 x i32] %arg4.coerce, i32 noundef %arg5)529void num_regs_left_64bit_aggregate(int4 arg0, int4 arg1, int4 arg2, int3 arg3, struct_char_x8 arg4, int arg5) { }530