1*532dc62bSNikita Popov// RUN: %clang_cc1 -no-opaque-pointers %s -triple spir-unknown-unknown -O0 -emit-llvm -o - | FileCheck -check-prefix=PRIVATE0 %s
2*532dc62bSNikita Popov// RUN: %clang_cc1 -no-opaque-pointers %s -triple amdgcn-amd-amdhsa-unknown -O0 -emit-llvm -o - | FileCheck -check-prefix=PRIVATE5 %s
3abdcfc18SAlexey Bader
4abdcfc18SAlexey Bader// CHECK: @test.arr = private unnamed_addr addrspace(2) constant [3 x i32] [i32 1, i32 2, i32 3], align 4
5abdcfc18SAlexey Bader
6abdcfc18SAlexey Badervoid test() {
7abdcfc18SAlexey Bader  __private int arr[] = {1, 2, 3};
8935574a4SMatt Arsenault// PRIVATE0:  %[[arr_i8_ptr:[0-9]+]] = bitcast [3 x i32]* %arr to i8*
93a881e6bSJF Bastien// PRIVATE0:  call void @llvm.memcpy.p0i8.p2i8.i32(i8* align 4 %[[arr_i8_ptr]], i8 addrspace(2)* align 4 bitcast ([3 x i32] addrspace(2)* @__const.test.arr to i8 addrspace(2)*), i32 12, i1 false)
10935574a4SMatt Arsenault
11935574a4SMatt Arsenault// PRIVATE5: %arr = alloca [3 x i32], align 4, addrspace(5)
12935574a4SMatt Arsenault// PRIVATE5: %0 = bitcast [3 x i32] addrspace(5)* %arr to i8 addrspace(5)*
133a881e6bSJF Bastien// PRIVATE5: call void @llvm.memcpy.p5i8.p4i8.i64(i8 addrspace(5)* align 4 %0, i8 addrspace(4)* align 4 bitcast ([3 x i32] addrspace(4)* @__const.test.arr to i8 addrspace(4)*), i64 12, i1 false)
14935574a4SMatt Arsenault}
15935574a4SMatt Arsenault
16935574a4SMatt Arsenault__kernel void initializer_cast_is_valid_crash() {
17935574a4SMatt Arsenault// PRIVATE0: %v512 = alloca [64 x i8], align 1
18935574a4SMatt Arsenault// PRIVATE0: %0 = bitcast [64 x i8]* %v512 to i8*
19935574a4SMatt Arsenault// PRIVATE0: call void @llvm.memset.p0i8.i32(i8* align 1 %0, i8 0, i32 64, i1 false)
20935574a4SMatt Arsenault// PRIVATE0: %1 = bitcast i8* %0 to [64 x i8]*
21935574a4SMatt Arsenault
22935574a4SMatt Arsenault
23935574a4SMatt Arsenault// PRIVATE5: %v512 = alloca [64 x i8], align 1, addrspace(5)
24935574a4SMatt Arsenault// PRIVATE5: %0 = bitcast [64 x i8] addrspace(5)* %v512 to i8 addrspace(5)*
25935574a4SMatt Arsenault// PRIVATE5: call void @llvm.memset.p5i8.i64(i8 addrspace(5)* align 1 %0, i8 0, i64 64, i1 false)
26935574a4SMatt Arsenault// PRIVATE5: %1 = bitcast i8 addrspace(5)* %0 to [64 x i8] addrspace(5)*
27935574a4SMatt Arsenault  unsigned char v512[64] = {
28935574a4SMatt Arsenault      0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,
29935574a4SMatt Arsenault      0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,
30935574a4SMatt Arsenault      0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,
31935574a4SMatt Arsenault      0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x00,0x02,0x00
32935574a4SMatt Arsenault  };
33abdcfc18SAlexey Bader}
34