brintos / llvm-project-archived public Read only

1// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --include-generated-funcs --replace-value-regex "__omp_offloading_[0-9a-z]+_[0-9a-z]+" "reduction_size[.].+[.]" "pl_cond[.].+[.|,]" --prefix-filecheck-ir-name _ --version 52// expected-no-diagnostics3 4// RUN: %clang_cc1 -no-enable-noundef-analysis  -verify -Wno-vla  -fopenmp -fopenmp-version=60 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  %s --check-prefix CK-645// RUN: %clang_cc1 -no-enable-noundef-analysis  -fopenmp -fopenmp-version=60 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s6// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp -fopenmp-version=60 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify -Wno-vla  %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  %s --check-prefix CK-647// RUN: %clang_cc1 -no-enable-noundef-analysis  -verify -Wno-vla  -fopenmp -fopenmp-version=60 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  %s --check-prefix CK-328// RUN: %clang_cc1 -no-enable-noundef-analysis  -fopenmp -fopenmp-version=60 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s9// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp -fopenmp-version=60 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify -Wno-vla  %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  %s --check-prefix CK-3210 11// RUN: %clang_cc1 -no-enable-noundef-analysis  -verify -Wno-vla  -fopenmp-simd -fopenmp-version=60 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  --check-prefix SIMD-ONLY-64 %s12// RUN: %clang_cc1 -no-enable-noundef-analysis  -fopenmp-simd -fopenmp-version=60 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s13// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp-simd -fopenmp-version=60 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify -Wno-vla  %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  --check-prefix SIMD-ONLY-64 %s14// RUN: %clang_cc1 -no-enable-noundef-analysis  -verify -Wno-vla  -fopenmp-simd -fopenmp-version=60 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  --check-prefix SIMD-ONLY-32 %s15// RUN: %clang_cc1 -no-enable-noundef-analysis  -fopenmp-simd -fopenmp-version=60 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s16// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp-simd -fopenmp-version=60 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify -Wno-vla  %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  --check-prefix SIMD-ONLY-32 %s17 18#ifndef HEADER19#define HEADER20void foo1(int a) {21  double d = (double)a;22 23  #pragma omp target default(private: scalar)24  {25    d += 1.0;26  }27}28 29void foo2() {30  int pvtArr[10];31 32  #pragma omp target default(private: aggregate)33  {34    pvtArr[5]++;35  }36}37 38void foo3() {39  int *pa;40 41  #pragma omp target default(private: pointer)42  {43    pa[50]++;44  }45}46 47// Specified variable-category doesn't apply to referenced variable, so48// normal implicitly determined data-sharing applies.49void foo4() {50  int p;51 52  #pragma omp target default(private: pointer)53  {54    p++;55  }56}57 58// Verify default clause with variable-category 'all' is equivalent to no59// variable-category. IR checks generated with 'all' but test runs without60// variable-category.61void foo5(int a) {62  double d = (double)a;63  int pvtArr[10];64  int *pa;65 66  #pragma omp target default(private)67  {68    d += 1.0;69    pvtArr[5]++;70    pa[50]++;71  }72}73 74// Verify default clause with 'shared' DSA is ignored. This makes it75// equivalent to target with no default clause. IR checks generated with76// no default clause but test runs with default 'shared'.77void foo6(int a) {78  double d = (double)a;79  int pvtArr[10];80  int *pa;81 82  #pragma omp target default(shared)83  {84    d += 1.0;85    pvtArr[5]++;86    pa[50]++;87  }88}89 90// Verify default clause with 'firstprivate' DSA is equivalent to specifying91// defaultmap with 'firstprivate'. IR checks generated with92// defaultmap(firstprivate) but test runs with default(firstprivate).93void foo7(int a) {94  double d = (double)a;95  int pvtArr[10];96  int *pa;97 98  #pragma omp target default(firstprivate)99  {100    d += 1.0;101    pvtArr[5]++;102    pa[50]++;103  }104}105 106// Verify 'default' clause on a combined 'target' directive is equivalent to107// specifying its constituent directives with 'default' clauses. IR checks108// generated with constituent directives but test runs with combined109// directive.110void foo8() {111  int x = 0;112  #pragma omp target teams distribute parallel for default(firstprivate) firstprivate(x)113  for (int i=0; i<10; i++)114    x += 1;115}116#endif // HEADER117// CK-64-LABEL: define dso_local void @_Z4foo1i(118// CK-64-SAME: i32 signext [[A:%.*]]) #[[ATTR0:[0-9]+]] {119// CK-64-NEXT:  [[ENTRY:.*:]]120// CK-64-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4121// CK-64-NEXT:    [[D:%.*]] = alloca double, align 8122// CK-64-NEXT:    [[D_CASTED:%.*]] = alloca i64, align 8123// CK-64-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 8124// CK-64-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 8125// CK-64-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 8126// CK-64-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8127// CK-64-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 4128// CK-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4129// CK-64-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double130// CK-64-NEXT:    store double [[CONV]], ptr [[D]], align 8131// CK-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D]], align 8132// CK-64-NEXT:    store double [[TMP1]], ptr [[D_CASTED]], align 8133// CK-64-NEXT:    [[TMP2:%.*]] = load i64, ptr [[D_CASTED]], align 8134// CK-64-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0135// CK-64-NEXT:    store i64 [[TMP2]], ptr [[TMP3]], align 8136// CK-64-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0137// CK-64-NEXT:    store i64 [[TMP2]], ptr [[TMP4]], align 8138// CK-64-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0139// CK-64-NEXT:    store ptr null, ptr [[TMP5]], align 8140// CK-64-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0141// CK-64-NEXT:    [[TMP7:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0142// CK-64-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0143// CK-64-NEXT:    store i32 3, ptr [[TMP8]], align 4144// CK-64-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1145// CK-64-NEXT:    store i32 1, ptr [[TMP9]], align 4146// CK-64-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2147// CK-64-NEXT:    store ptr [[TMP6]], ptr [[TMP10]], align 8148// CK-64-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3149// CK-64-NEXT:    store ptr [[TMP7]], ptr [[TMP11]], align 8150// CK-64-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4151// CK-64-NEXT:    store ptr @.offload_sizes, ptr [[TMP12]], align 8152// CK-64-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5153// CK-64-NEXT:    store ptr @.offload_maptypes, ptr [[TMP13]], align 8154// CK-64-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6155// CK-64-NEXT:    store ptr null, ptr [[TMP14]], align 8156// CK-64-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7157// CK-64-NEXT:    store ptr null, ptr [[TMP15]], align 8158// CK-64-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8159// CK-64-NEXT:    store i64 0, ptr [[TMP16]], align 8160// CK-64-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9161// CK-64-NEXT:    store i64 0, ptr [[TMP17]], align 8162// CK-64-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10163// CK-64-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP18]], align 4164// CK-64-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11165// CK-64-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP19]], align 4166// CK-64-NEXT:    [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12167// CK-64-NEXT:    store i32 0, ptr [[TMP20]], align 4168// CK-64-NEXT:    [[TMP21:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1:[0-9]+]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo1i_l23.region_id, ptr [[KERNEL_ARGS]])169// CK-64-NEXT:    [[TMP22:%.*]] = icmp ne i32 [[TMP21]], 0170// CK-64-NEXT:    br i1 [[TMP22]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]171// CK-64:       [[OMP_OFFLOAD_FAILED]]:172// CK-64-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo1i_l23(i64 [[TMP2]]) #[[ATTR2:[0-9]+]]173// CK-64-NEXT:    br label %[[OMP_OFFLOAD_CONT]]174// CK-64:       [[OMP_OFFLOAD_CONT]]:175// CK-64-NEXT:    ret void176//177//178// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo1i_l23(179// CK-64-SAME: i64 [[D:%.*]]) #[[ATTR1:[0-9]+]] {180// CK-64-NEXT:  [[ENTRY:.*:]]181// CK-64-NEXT:    [[D_ADDR:%.*]] = alloca i64, align 8182// CK-64-NEXT:    [[D1:%.*]] = alloca double, align 8183// CK-64-NEXT:    store i64 [[D]], ptr [[D_ADDR]], align 8184// CK-64-NEXT:    [[TMP0:%.*]] = load double, ptr [[D1]], align 8185// CK-64-NEXT:    [[ADD:%.*]] = fadd double [[TMP0]], 1.000000e+00186// CK-64-NEXT:    store double [[ADD]], ptr [[D1]], align 8187// CK-64-NEXT:    ret void188//189//190// CK-64-LABEL: define dso_local void @_Z4foo2v(191// CK-64-SAME: ) #[[ATTR0]] {192// CK-64-NEXT:  [[ENTRY:.*:]]193// CK-64-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 4194// CK-64-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 8195// CK-64-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 8196// CK-64-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 8197// CK-64-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8198// CK-64-NEXT:    [[TMP0:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0199// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[TMP0]], align 8200// CK-64-NEXT:    [[TMP1:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0201// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[TMP1]], align 8202// CK-64-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0203// CK-64-NEXT:    store ptr null, ptr [[TMP2]], align 8204// CK-64-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0205// CK-64-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0206// CK-64-NEXT:    [[TMP5:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0207// CK-64-NEXT:    store i32 3, ptr [[TMP5]], align 4208// CK-64-NEXT:    [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1209// CK-64-NEXT:    store i32 1, ptr [[TMP6]], align 4210// CK-64-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2211// CK-64-NEXT:    store ptr [[TMP3]], ptr [[TMP7]], align 8212// CK-64-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3213// CK-64-NEXT:    store ptr [[TMP4]], ptr [[TMP8]], align 8214// CK-64-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4215// CK-64-NEXT:    store ptr @.offload_sizes.1, ptr [[TMP9]], align 8216// CK-64-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5217// CK-64-NEXT:    store ptr @.offload_maptypes.2, ptr [[TMP10]], align 8218// CK-64-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6219// CK-64-NEXT:    store ptr null, ptr [[TMP11]], align 8220// CK-64-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7221// CK-64-NEXT:    store ptr null, ptr [[TMP12]], align 8222// CK-64-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8223// CK-64-NEXT:    store i64 0, ptr [[TMP13]], align 8224// CK-64-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9225// CK-64-NEXT:    store i64 0, ptr [[TMP14]], align 8226// CK-64-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10227// CK-64-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP15]], align 4228// CK-64-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11229// CK-64-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP16]], align 4230// CK-64-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12231// CK-64-NEXT:    store i32 0, ptr [[TMP17]], align 4232// CK-64-NEXT:    [[TMP18:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo2v_l32.region_id, ptr [[KERNEL_ARGS]])233// CK-64-NEXT:    [[TMP19:%.*]] = icmp ne i32 [[TMP18]], 0234// CK-64-NEXT:    br i1 [[TMP19]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]235// CK-64:       [[OMP_OFFLOAD_FAILED]]:236// CK-64-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo2v_l32(ptr [[PVTARR]]) #[[ATTR2]]237// CK-64-NEXT:    br label %[[OMP_OFFLOAD_CONT]]238// CK-64:       [[OMP_OFFLOAD_CONT]]:239// CK-64-NEXT:    ret void240//241//242// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo2v_l32(243// CK-64-SAME: ptr nonnull align 4 dereferenceable(40) [[PVTARR:%.*]]) #[[ATTR1]] {244// CK-64-NEXT:  [[ENTRY:.*:]]245// CK-64-NEXT:    [[PVTARR_ADDR:%.*]] = alloca ptr, align 8246// CK-64-NEXT:    [[PVTARR1:%.*]] = alloca [10 x i32], align 4247// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[PVTARR_ADDR]], align 8248// CK-64-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PVTARR_ADDR]], align 8, !nonnull [[META19:![0-9]+]], !align [[META20:![0-9]+]]249// CK-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR1]], i64 0, i64 5250// CK-64-NEXT:    [[TMP1:%.*]] = load i32, ptr [[ARRAYIDX]], align 4251// CK-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP1]], 1252// CK-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 4253// CK-64-NEXT:    ret void254//255//256// CK-64-LABEL: define dso_local void @_Z4foo3v(257// CK-64-SAME: ) #[[ATTR0]] {258// CK-64-NEXT:  [[ENTRY:.*:]]259// CK-64-NEXT:    [[PA:%.*]] = alloca ptr, align 8260// CK-64-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 8261// CK-64-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 8262// CK-64-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 8263// CK-64-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8264// CK-64-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PA]], align 8265// CK-64-NEXT:    [[TMP1:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0266// CK-64-NEXT:    store ptr [[TMP0]], ptr [[TMP1]], align 8267// CK-64-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0268// CK-64-NEXT:    store ptr [[TMP0]], ptr [[TMP2]], align 8269// CK-64-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0270// CK-64-NEXT:    store ptr null, ptr [[TMP3]], align 8271// CK-64-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0272// CK-64-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0273// CK-64-NEXT:    [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0274// CK-64-NEXT:    store i32 3, ptr [[TMP6]], align 4275// CK-64-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1276// CK-64-NEXT:    store i32 1, ptr [[TMP7]], align 4277// CK-64-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2278// CK-64-NEXT:    store ptr [[TMP4]], ptr [[TMP8]], align 8279// CK-64-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3280// CK-64-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 8281// CK-64-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4282// CK-64-NEXT:    store ptr @.offload_sizes.3, ptr [[TMP10]], align 8283// CK-64-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5284// CK-64-NEXT:    store ptr @.offload_maptypes.4, ptr [[TMP11]], align 8285// CK-64-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6286// CK-64-NEXT:    store ptr null, ptr [[TMP12]], align 8287// CK-64-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7288// CK-64-NEXT:    store ptr null, ptr [[TMP13]], align 8289// CK-64-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8290// CK-64-NEXT:    store i64 0, ptr [[TMP14]], align 8291// CK-64-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9292// CK-64-NEXT:    store i64 0, ptr [[TMP15]], align 8293// CK-64-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10294// CK-64-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP16]], align 4295// CK-64-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11296// CK-64-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP17]], align 4297// CK-64-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12298// CK-64-NEXT:    store i32 0, ptr [[TMP18]], align 4299// CK-64-NEXT:    [[TMP19:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo3v_l41.region_id, ptr [[KERNEL_ARGS]])300// CK-64-NEXT:    [[TMP20:%.*]] = icmp ne i32 [[TMP19]], 0301// CK-64-NEXT:    br i1 [[TMP20]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]302// CK-64:       [[OMP_OFFLOAD_FAILED]]:303// CK-64-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo3v_l41(ptr [[TMP0]]) #[[ATTR2]]304// CK-64-NEXT:    br label %[[OMP_OFFLOAD_CONT]]305// CK-64:       [[OMP_OFFLOAD_CONT]]:306// CK-64-NEXT:    ret void307//308//309// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo3v_l41(310// CK-64-SAME: ptr [[PA:%.*]]) #[[ATTR1]] {311// CK-64-NEXT:  [[ENTRY:.*:]]312// CK-64-NEXT:    [[PA_ADDR:%.*]] = alloca ptr, align 8313// CK-64-NEXT:    [[PA1:%.*]] = alloca ptr, align 8314// CK-64-NEXT:    store ptr [[PA]], ptr [[PA_ADDR]], align 8315// CK-64-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PA1]], align 8316// CK-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds i32, ptr [[TMP0]], i64 50317// CK-64-NEXT:    [[TMP1:%.*]] = load i32, ptr [[ARRAYIDX]], align 4318// CK-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP1]], 1319// CK-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 4320// CK-64-NEXT:    ret void321//322//323// CK-64-LABEL: define dso_local void @_Z4foo4v(324// CK-64-SAME: ) #[[ATTR0]] {325// CK-64-NEXT:  [[ENTRY:.*:]]326// CK-64-NEXT:    [[P:%.*]] = alloca i32, align 4327// CK-64-NEXT:    [[P_CASTED:%.*]] = alloca i64, align 8328// CK-64-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 8329// CK-64-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 8330// CK-64-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 8331// CK-64-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8332// CK-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[P]], align 4333// CK-64-NEXT:    store i32 [[TMP0]], ptr [[P_CASTED]], align 4334// CK-64-NEXT:    [[TMP1:%.*]] = load i64, ptr [[P_CASTED]], align 8335// CK-64-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0336// CK-64-NEXT:    store i64 [[TMP1]], ptr [[TMP2]], align 8337// CK-64-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0338// CK-64-NEXT:    store i64 [[TMP1]], ptr [[TMP3]], align 8339// CK-64-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0340// CK-64-NEXT:    store ptr null, ptr [[TMP4]], align 8341// CK-64-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0342// CK-64-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0343// CK-64-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0344// CK-64-NEXT:    store i32 3, ptr [[TMP7]], align 4345// CK-64-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1346// CK-64-NEXT:    store i32 1, ptr [[TMP8]], align 4347// CK-64-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2348// CK-64-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 8349// CK-64-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3350// CK-64-NEXT:    store ptr [[TMP6]], ptr [[TMP10]], align 8351// CK-64-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4352// CK-64-NEXT:    store ptr @.offload_sizes.5, ptr [[TMP11]], align 8353// CK-64-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5354// CK-64-NEXT:    store ptr @.offload_maptypes.6, ptr [[TMP12]], align 8355// CK-64-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6356// CK-64-NEXT:    store ptr null, ptr [[TMP13]], align 8357// CK-64-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7358// CK-64-NEXT:    store ptr null, ptr [[TMP14]], align 8359// CK-64-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8360// CK-64-NEXT:    store i64 0, ptr [[TMP15]], align 8361// CK-64-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9362// CK-64-NEXT:    store i64 0, ptr [[TMP16]], align 8363// CK-64-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10364// CK-64-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP17]], align 4365// CK-64-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11366// CK-64-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP18]], align 4367// CK-64-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12368// CK-64-NEXT:    store i32 0, ptr [[TMP19]], align 4369// CK-64-NEXT:    [[TMP20:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo4v_l52.region_id, ptr [[KERNEL_ARGS]])370// CK-64-NEXT:    [[TMP21:%.*]] = icmp ne i32 [[TMP20]], 0371// CK-64-NEXT:    br i1 [[TMP21]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]372// CK-64:       [[OMP_OFFLOAD_FAILED]]:373// CK-64-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo4v_l52(i64 [[TMP1]]) #[[ATTR2]]374// CK-64-NEXT:    br label %[[OMP_OFFLOAD_CONT]]375// CK-64:       [[OMP_OFFLOAD_CONT]]:376// CK-64-NEXT:    ret void377//378//379// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo4v_l52(380// CK-64-SAME: i64 [[P:%.*]]) #[[ATTR1]] {381// CK-64-NEXT:  [[ENTRY:.*:]]382// CK-64-NEXT:    [[P_ADDR:%.*]] = alloca i64, align 8383// CK-64-NEXT:    store i64 [[P]], ptr [[P_ADDR]], align 8384// CK-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[P_ADDR]], align 4385// CK-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP0]], 1386// CK-64-NEXT:    store i32 [[INC]], ptr [[P_ADDR]], align 4387// CK-64-NEXT:    ret void388//389//390// CK-64-LABEL: define dso_local void @_Z4foo5i(391// CK-64-SAME: i32 signext [[A:%.*]]) #[[ATTR0]] {392// CK-64-NEXT:  [[ENTRY:.*:]]393// CK-64-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4394// CK-64-NEXT:    [[D:%.*]] = alloca double, align 8395// CK-64-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 4396// CK-64-NEXT:    [[PA:%.*]] = alloca ptr, align 8397// CK-64-NEXT:    [[D_CASTED:%.*]] = alloca i64, align 8398// CK-64-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x ptr], align 8399// CK-64-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x ptr], align 8400// CK-64-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x ptr], align 8401// CK-64-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8402// CK-64-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 4403// CK-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4404// CK-64-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double405// CK-64-NEXT:    store double [[CONV]], ptr [[D]], align 8406// CK-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D]], align 8407// CK-64-NEXT:    store double [[TMP1]], ptr [[D_CASTED]], align 8408// CK-64-NEXT:    [[TMP2:%.*]] = load i64, ptr [[D_CASTED]], align 8409// CK-64-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA]], align 8410// CK-64-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0411// CK-64-NEXT:    store i64 [[TMP2]], ptr [[TMP4]], align 8412// CK-64-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0413// CK-64-NEXT:    store i64 [[TMP2]], ptr [[TMP5]], align 8414// CK-64-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0415// CK-64-NEXT:    store ptr null, ptr [[TMP6]], align 8416// CK-64-NEXT:    [[TMP7:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1417// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[TMP7]], align 8418// CK-64-NEXT:    [[TMP8:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1419// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[TMP8]], align 8420// CK-64-NEXT:    [[TMP9:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1421// CK-64-NEXT:    store ptr null, ptr [[TMP9]], align 8422// CK-64-NEXT:    [[TMP10:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2423// CK-64-NEXT:    store ptr [[TMP3]], ptr [[TMP10]], align 8424// CK-64-NEXT:    [[TMP11:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 2425// CK-64-NEXT:    store ptr [[TMP3]], ptr [[TMP11]], align 8426// CK-64-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2427// CK-64-NEXT:    store ptr null, ptr [[TMP12]], align 8428// CK-64-NEXT:    [[TMP13:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0429// CK-64-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0430// CK-64-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0431// CK-64-NEXT:    store i32 3, ptr [[TMP15]], align 4432// CK-64-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1433// CK-64-NEXT:    store i32 3, ptr [[TMP16]], align 4434// CK-64-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2435// CK-64-NEXT:    store ptr [[TMP13]], ptr [[TMP17]], align 8436// CK-64-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3437// CK-64-NEXT:    store ptr [[TMP14]], ptr [[TMP18]], align 8438// CK-64-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4439// CK-64-NEXT:    store ptr @.offload_sizes.7, ptr [[TMP19]], align 8440// CK-64-NEXT:    [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5441// CK-64-NEXT:    store ptr @.offload_maptypes.8, ptr [[TMP20]], align 8442// CK-64-NEXT:    [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6443// CK-64-NEXT:    store ptr null, ptr [[TMP21]], align 8444// CK-64-NEXT:    [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7445// CK-64-NEXT:    store ptr null, ptr [[TMP22]], align 8446// CK-64-NEXT:    [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8447// CK-64-NEXT:    store i64 0, ptr [[TMP23]], align 8448// CK-64-NEXT:    [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9449// CK-64-NEXT:    store i64 0, ptr [[TMP24]], align 8450// CK-64-NEXT:    [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10451// CK-64-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP25]], align 4452// CK-64-NEXT:    [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11453// CK-64-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP26]], align 4454// CK-64-NEXT:    [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12455// CK-64-NEXT:    store i32 0, ptr [[TMP27]], align 4456// CK-64-NEXT:    [[TMP28:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo5i_l66.region_id, ptr [[KERNEL_ARGS]])457// CK-64-NEXT:    [[TMP29:%.*]] = icmp ne i32 [[TMP28]], 0458// CK-64-NEXT:    br i1 [[TMP29]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]459// CK-64:       [[OMP_OFFLOAD_FAILED]]:460// CK-64-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo5i_l66(i64 [[TMP2]], ptr [[PVTARR]], ptr [[TMP3]]) #[[ATTR2]]461// CK-64-NEXT:    br label %[[OMP_OFFLOAD_CONT]]462// CK-64:       [[OMP_OFFLOAD_CONT]]:463// CK-64-NEXT:    ret void464//465//466// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo5i_l66(467// CK-64-SAME: i64 [[D:%.*]], ptr nonnull align 4 dereferenceable(40) [[PVTARR:%.*]], ptr [[PA:%.*]]) #[[ATTR1]] {468// CK-64-NEXT:  [[ENTRY:.*:]]469// CK-64-NEXT:    [[D_ADDR:%.*]] = alloca i64, align 8470// CK-64-NEXT:    [[PVTARR_ADDR:%.*]] = alloca ptr, align 8471// CK-64-NEXT:    [[PA_ADDR:%.*]] = alloca ptr, align 8472// CK-64-NEXT:    [[D1:%.*]] = alloca double, align 8473// CK-64-NEXT:    [[PVTARR2:%.*]] = alloca [10 x i32], align 4474// CK-64-NEXT:    [[PA3:%.*]] = alloca ptr, align 8475// CK-64-NEXT:    store i64 [[D]], ptr [[D_ADDR]], align 8476// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[PVTARR_ADDR]], align 8477// CK-64-NEXT:    store ptr [[PA]], ptr [[PA_ADDR]], align 8478// CK-64-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PVTARR_ADDR]], align 8, !nonnull [[META19]], !align [[META20]]479// CK-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D1]], align 8480// CK-64-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+00481// CK-64-NEXT:    store double [[ADD]], ptr [[D1]], align 8482// CK-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR2]], i64 0, i64 5483// CK-64-NEXT:    [[TMP2:%.*]] = load i32, ptr [[ARRAYIDX]], align 4484// CK-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 1485// CK-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 4486// CK-64-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA3]], align 8487// CK-64-NEXT:    [[ARRAYIDX4:%.*]] = getelementptr inbounds i32, ptr [[TMP3]], i64 50488// CK-64-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX4]], align 4489// CK-64-NEXT:    [[INC5:%.*]] = add nsw i32 [[TMP4]], 1490// CK-64-NEXT:    store i32 [[INC5]], ptr [[ARRAYIDX4]], align 4491// CK-64-NEXT:    ret void492//493//494// CK-64-LABEL: define dso_local void @_Z4foo6i(495// CK-64-SAME: i32 signext [[A:%.*]]) #[[ATTR0]] {496// CK-64-NEXT:  [[ENTRY:.*:]]497// CK-64-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4498// CK-64-NEXT:    [[D:%.*]] = alloca double, align 8499// CK-64-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 4500// CK-64-NEXT:    [[PA:%.*]] = alloca ptr, align 8501// CK-64-NEXT:    [[D_CASTED:%.*]] = alloca i64, align 8502// CK-64-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x ptr], align 8503// CK-64-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x ptr], align 8504// CK-64-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x ptr], align 8505// CK-64-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8506// CK-64-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 4507// CK-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4508// CK-64-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double509// CK-64-NEXT:    store double [[CONV]], ptr [[D]], align 8510// CK-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D]], align 8511// CK-64-NEXT:    store double [[TMP1]], ptr [[D_CASTED]], align 8512// CK-64-NEXT:    [[TMP2:%.*]] = load i64, ptr [[D_CASTED]], align 8513// CK-64-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA]], align 8514// CK-64-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0515// CK-64-NEXT:    store i64 [[TMP2]], ptr [[TMP4]], align 8516// CK-64-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0517// CK-64-NEXT:    store i64 [[TMP2]], ptr [[TMP5]], align 8518// CK-64-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0519// CK-64-NEXT:    store ptr null, ptr [[TMP6]], align 8520// CK-64-NEXT:    [[TMP7:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1521// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[TMP7]], align 8522// CK-64-NEXT:    [[TMP8:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1523// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[TMP8]], align 8524// CK-64-NEXT:    [[TMP9:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1525// CK-64-NEXT:    store ptr null, ptr [[TMP9]], align 8526// CK-64-NEXT:    [[TMP10:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2527// CK-64-NEXT:    store ptr [[TMP3]], ptr [[TMP10]], align 8528// CK-64-NEXT:    [[TMP11:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 2529// CK-64-NEXT:    store ptr [[TMP3]], ptr [[TMP11]], align 8530// CK-64-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2531// CK-64-NEXT:    store ptr null, ptr [[TMP12]], align 8532// CK-64-NEXT:    [[TMP13:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0533// CK-64-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0534// CK-64-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0535// CK-64-NEXT:    store i32 3, ptr [[TMP15]], align 4536// CK-64-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1537// CK-64-NEXT:    store i32 3, ptr [[TMP16]], align 4538// CK-64-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2539// CK-64-NEXT:    store ptr [[TMP13]], ptr [[TMP17]], align 8540// CK-64-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3541// CK-64-NEXT:    store ptr [[TMP14]], ptr [[TMP18]], align 8542// CK-64-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4543// CK-64-NEXT:    store ptr @.offload_sizes.9, ptr [[TMP19]], align 8544// CK-64-NEXT:    [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5545// CK-64-NEXT:    store ptr @.offload_maptypes.10, ptr [[TMP20]], align 8546// CK-64-NEXT:    [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6547// CK-64-NEXT:    store ptr null, ptr [[TMP21]], align 8548// CK-64-NEXT:    [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7549// CK-64-NEXT:    store ptr null, ptr [[TMP22]], align 8550// CK-64-NEXT:    [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8551// CK-64-NEXT:    store i64 0, ptr [[TMP23]], align 8552// CK-64-NEXT:    [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9553// CK-64-NEXT:    store i64 0, ptr [[TMP24]], align 8554// CK-64-NEXT:    [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10555// CK-64-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP25]], align 4556// CK-64-NEXT:    [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11557// CK-64-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP26]], align 4558// CK-64-NEXT:    [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12559// CK-64-NEXT:    store i32 0, ptr [[TMP27]], align 4560// CK-64-NEXT:    [[TMP28:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo6i_l82.region_id, ptr [[KERNEL_ARGS]])561// CK-64-NEXT:    [[TMP29:%.*]] = icmp ne i32 [[TMP28]], 0562// CK-64-NEXT:    br i1 [[TMP29]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]563// CK-64:       [[OMP_OFFLOAD_FAILED]]:564// CK-64-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo6i_l82(i64 [[TMP2]], ptr [[PVTARR]], ptr [[TMP3]]) #[[ATTR2]]565// CK-64-NEXT:    br label %[[OMP_OFFLOAD_CONT]]566// CK-64:       [[OMP_OFFLOAD_CONT]]:567// CK-64-NEXT:    ret void568//569//570// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo6i_l82(571// CK-64-SAME: i64 [[D:%.*]], ptr nonnull align 4 dereferenceable(40) [[PVTARR:%.*]], ptr [[PA:%.*]]) #[[ATTR1]] {572// CK-64-NEXT:  [[ENTRY:.*:]]573// CK-64-NEXT:    [[D_ADDR:%.*]] = alloca i64, align 8574// CK-64-NEXT:    [[PVTARR_ADDR:%.*]] = alloca ptr, align 8575// CK-64-NEXT:    [[PA_ADDR:%.*]] = alloca ptr, align 8576// CK-64-NEXT:    store i64 [[D]], ptr [[D_ADDR]], align 8577// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[PVTARR_ADDR]], align 8578// CK-64-NEXT:    store ptr [[PA]], ptr [[PA_ADDR]], align 8579// CK-64-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PVTARR_ADDR]], align 8, !nonnull [[META19]], !align [[META20]]580// CK-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D_ADDR]], align 8581// CK-64-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+00582// CK-64-NEXT:    store double [[ADD]], ptr [[D_ADDR]], align 8583// CK-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[TMP0]], i64 0, i64 5584// CK-64-NEXT:    [[TMP2:%.*]] = load i32, ptr [[ARRAYIDX]], align 4585// CK-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 1586// CK-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 4587// CK-64-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA_ADDR]], align 8588// CK-64-NEXT:    [[ARRAYIDX1:%.*]] = getelementptr inbounds i32, ptr [[TMP3]], i64 50589// CK-64-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX1]], align 4590// CK-64-NEXT:    [[INC2:%.*]] = add nsw i32 [[TMP4]], 1591// CK-64-NEXT:    store i32 [[INC2]], ptr [[ARRAYIDX1]], align 4592// CK-64-NEXT:    ret void593//594//595// CK-64-LABEL: define dso_local void @_Z4foo7i(596// CK-64-SAME: i32 signext [[A:%.*]]) #[[ATTR0]] {597// CK-64-NEXT:  [[ENTRY:.*:]]598// CK-64-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4599// CK-64-NEXT:    [[D:%.*]] = alloca double, align 8600// CK-64-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 4601// CK-64-NEXT:    [[PA:%.*]] = alloca ptr, align 8602// CK-64-NEXT:    [[D_CASTED:%.*]] = alloca i64, align 8603// CK-64-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x ptr], align 8604// CK-64-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x ptr], align 8605// CK-64-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x ptr], align 8606// CK-64-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8607// CK-64-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 4608// CK-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4609// CK-64-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double610// CK-64-NEXT:    store double [[CONV]], ptr [[D]], align 8611// CK-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D]], align 8612// CK-64-NEXT:    store double [[TMP1]], ptr [[D_CASTED]], align 8613// CK-64-NEXT:    [[TMP2:%.*]] = load i64, ptr [[D_CASTED]], align 8614// CK-64-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA]], align 8615// CK-64-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0616// CK-64-NEXT:    store i64 [[TMP2]], ptr [[TMP4]], align 8617// CK-64-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0618// CK-64-NEXT:    store i64 [[TMP2]], ptr [[TMP5]], align 8619// CK-64-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0620// CK-64-NEXT:    store ptr null, ptr [[TMP6]], align 8621// CK-64-NEXT:    [[TMP7:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1622// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[TMP7]], align 8623// CK-64-NEXT:    [[TMP8:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1624// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[TMP8]], align 8625// CK-64-NEXT:    [[TMP9:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1626// CK-64-NEXT:    store ptr null, ptr [[TMP9]], align 8627// CK-64-NEXT:    [[TMP10:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2628// CK-64-NEXT:    store ptr [[TMP3]], ptr [[TMP10]], align 8629// CK-64-NEXT:    [[TMP11:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 2630// CK-64-NEXT:    store ptr [[TMP3]], ptr [[TMP11]], align 8631// CK-64-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2632// CK-64-NEXT:    store ptr null, ptr [[TMP12]], align 8633// CK-64-NEXT:    [[TMP13:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0634// CK-64-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0635// CK-64-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0636// CK-64-NEXT:    store i32 3, ptr [[TMP15]], align 4637// CK-64-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1638// CK-64-NEXT:    store i32 3, ptr [[TMP16]], align 4639// CK-64-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2640// CK-64-NEXT:    store ptr [[TMP13]], ptr [[TMP17]], align 8641// CK-64-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3642// CK-64-NEXT:    store ptr [[TMP14]], ptr [[TMP18]], align 8643// CK-64-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4644// CK-64-NEXT:    store ptr @.offload_sizes.11, ptr [[TMP19]], align 8645// CK-64-NEXT:    [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5646// CK-64-NEXT:    store ptr @.offload_maptypes.12, ptr [[TMP20]], align 8647// CK-64-NEXT:    [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6648// CK-64-NEXT:    store ptr null, ptr [[TMP21]], align 8649// CK-64-NEXT:    [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7650// CK-64-NEXT:    store ptr null, ptr [[TMP22]], align 8651// CK-64-NEXT:    [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8652// CK-64-NEXT:    store i64 0, ptr [[TMP23]], align 8653// CK-64-NEXT:    [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9654// CK-64-NEXT:    store i64 0, ptr [[TMP24]], align 8655// CK-64-NEXT:    [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10656// CK-64-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP25]], align 4657// CK-64-NEXT:    [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11658// CK-64-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP26]], align 4659// CK-64-NEXT:    [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12660// CK-64-NEXT:    store i32 0, ptr [[TMP27]], align 4661// CK-64-NEXT:    [[TMP28:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo7i_l98.region_id, ptr [[KERNEL_ARGS]])662// CK-64-NEXT:    [[TMP29:%.*]] = icmp ne i32 [[TMP28]], 0663// CK-64-NEXT:    br i1 [[TMP29]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]664// CK-64:       [[OMP_OFFLOAD_FAILED]]:665// CK-64-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo7i_l98(i64 [[TMP2]], ptr [[PVTARR]], ptr [[TMP3]]) #[[ATTR2]]666// CK-64-NEXT:    br label %[[OMP_OFFLOAD_CONT]]667// CK-64:       [[OMP_OFFLOAD_CONT]]:668// CK-64-NEXT:    ret void669//670//671// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo7i_l98(672// CK-64-SAME: i64 [[D:%.*]], ptr nonnull align 4 dereferenceable(40) [[PVTARR:%.*]], ptr [[PA:%.*]]) #[[ATTR1]] {673// CK-64-NEXT:  [[ENTRY:.*:]]674// CK-64-NEXT:    [[D_ADDR:%.*]] = alloca i64, align 8675// CK-64-NEXT:    [[PVTARR_ADDR:%.*]] = alloca ptr, align 8676// CK-64-NEXT:    [[PA_ADDR:%.*]] = alloca ptr, align 8677// CK-64-NEXT:    [[PVTARR1:%.*]] = alloca [10 x i32], align 4678// CK-64-NEXT:    store i64 [[D]], ptr [[D_ADDR]], align 8679// CK-64-NEXT:    store ptr [[PVTARR]], ptr [[PVTARR_ADDR]], align 8680// CK-64-NEXT:    store ptr [[PA]], ptr [[PA_ADDR]], align 8681// CK-64-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PVTARR_ADDR]], align 8, !nonnull [[META19]], !align [[META20]]682// CK-64-NEXT:    call void @llvm.memcpy.p0.p0.i64(ptr align 4 [[PVTARR1]], ptr align 4 [[TMP0]], i64 40, i1 false)683// CK-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D_ADDR]], align 8684// CK-64-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+00685// CK-64-NEXT:    store double [[ADD]], ptr [[D_ADDR]], align 8686// CK-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR1]], i64 0, i64 5687// CK-64-NEXT:    [[TMP2:%.*]] = load i32, ptr [[ARRAYIDX]], align 4688// CK-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 1689// CK-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 4690// CK-64-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA_ADDR]], align 8691// CK-64-NEXT:    [[ARRAYIDX2:%.*]] = getelementptr inbounds i32, ptr [[TMP3]], i64 50692// CK-64-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX2]], align 4693// CK-64-NEXT:    [[INC3:%.*]] = add nsw i32 [[TMP4]], 1694// CK-64-NEXT:    store i32 [[INC3]], ptr [[ARRAYIDX2]], align 4695// CK-64-NEXT:    ret void696//697//698// CK-64-LABEL: define dso_local void @_Z4foo8v(699// CK-64-SAME: ) #[[ATTR0]] {700// CK-64-NEXT:  [[ENTRY:.*:]]701// CK-64-NEXT:    [[X:%.*]] = alloca i32, align 4702// CK-64-NEXT:    [[X_CASTED:%.*]] = alloca i64, align 8703// CK-64-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 8704// CK-64-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 8705// CK-64-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 8706// CK-64-NEXT:    [[TMP:%.*]] = alloca i32, align 4707// CK-64-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8708// CK-64-NEXT:    store i32 0, ptr [[X]], align 4709// CK-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[X]], align 4710// CK-64-NEXT:    store i32 [[TMP0]], ptr [[X_CASTED]], align 4711// CK-64-NEXT:    [[TMP1:%.*]] = load i64, ptr [[X_CASTED]], align 8712// CK-64-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0713// CK-64-NEXT:    store i64 [[TMP1]], ptr [[TMP2]], align 8714// CK-64-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0715// CK-64-NEXT:    store i64 [[TMP1]], ptr [[TMP3]], align 8716// CK-64-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0717// CK-64-NEXT:    store ptr null, ptr [[TMP4]], align 8718// CK-64-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0719// CK-64-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0720// CK-64-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0721// CK-64-NEXT:    store i32 3, ptr [[TMP7]], align 4722// CK-64-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1723// CK-64-NEXT:    store i32 1, ptr [[TMP8]], align 4724// CK-64-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2725// CK-64-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 8726// CK-64-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3727// CK-64-NEXT:    store ptr [[TMP6]], ptr [[TMP10]], align 8728// CK-64-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4729// CK-64-NEXT:    store ptr @.offload_sizes.13, ptr [[TMP11]], align 8730// CK-64-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5731// CK-64-NEXT:    store ptr @.offload_maptypes.14, ptr [[TMP12]], align 8732// CK-64-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6733// CK-64-NEXT:    store ptr null, ptr [[TMP13]], align 8734// CK-64-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7735// CK-64-NEXT:    store ptr null, ptr [[TMP14]], align 8736// CK-64-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8737// CK-64-NEXT:    store i64 10, ptr [[TMP15]], align 8738// CK-64-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9739// CK-64-NEXT:    store i64 0, ptr [[TMP16]], align 8740// CK-64-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10741// CK-64-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP17]], align 4742// CK-64-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11743// CK-64-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP18]], align 4744// CK-64-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12745// CK-64-NEXT:    store i32 0, ptr [[TMP19]], align 4746// CK-64-NEXT:    [[TMP20:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.region_id, ptr [[KERNEL_ARGS]])747// CK-64-NEXT:    [[TMP21:%.*]] = icmp ne i32 [[TMP20]], 0748// CK-64-NEXT:    br i1 [[TMP21]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]749// CK-64:       [[OMP_OFFLOAD_FAILED]]:750// CK-64-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112(i64 [[TMP1]]) #[[ATTR2]]751// CK-64-NEXT:    br label %[[OMP_OFFLOAD_CONT]]752// CK-64:       [[OMP_OFFLOAD_CONT]]:753// CK-64-NEXT:    ret void754//755//756// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112(757// CK-64-SAME: i64 [[X:%.*]]) #[[ATTR1]] {758// CK-64-NEXT:  [[ENTRY:.*:]]759// CK-64-NEXT:    [[X_ADDR:%.*]] = alloca i64, align 8760// CK-64-NEXT:    [[X_CASTED:%.*]] = alloca i64, align 8761// CK-64-NEXT:    store i64 [[X]], ptr [[X_ADDR]], align 8762// CK-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[X_ADDR]], align 4763// CK-64-NEXT:    store i32 [[TMP0]], ptr [[X_CASTED]], align 4764// CK-64-NEXT:    [[TMP1:%.*]] = load i64, ptr [[X_CASTED]], align 8765// CK-64-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.omp_outlined, i64 [[TMP1]])766// CK-64-NEXT:    ret void767//768//769// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.omp_outlined(770// CK-64-SAME: ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i64 [[X:%.*]]) #[[ATTR1]] {771// CK-64-NEXT:  [[ENTRY:.*:]]772// CK-64-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8773// CK-64-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8774// CK-64-NEXT:    [[X_ADDR:%.*]] = alloca i64, align 8775// CK-64-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4776// CK-64-NEXT:    [[TMP:%.*]] = alloca i32, align 4777// CK-64-NEXT:    [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4778// CK-64-NEXT:    [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4779// CK-64-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4780// CK-64-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4781// CK-64-NEXT:    [[I:%.*]] = alloca i32, align 4782// CK-64-NEXT:    [[X_CASTED:%.*]] = alloca i64, align 8783// CK-64-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8784// CK-64-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8785// CK-64-NEXT:    store i64 [[X]], ptr [[X_ADDR]], align 8786// CK-64-NEXT:    store i32 0, ptr [[DOTOMP_COMB_LB]], align 4787// CK-64-NEXT:    store i32 9, ptr [[DOTOMP_COMB_UB]], align 4788// CK-64-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4789// CK-64-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4790// CK-64-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8791// CK-64-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4792// CK-64-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB2:[0-9]+]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)793// CK-64-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4794// CK-64-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 9795// CK-64-NEXT:    br i1 [[CMP]], label %[[COND_TRUE:.*]], label %[[COND_FALSE:.*]]796// CK-64:       [[COND_TRUE]]:797// CK-64-NEXT:    br label %[[COND_END:.*]]798// CK-64:       [[COND_FALSE]]:799// CK-64-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4800// CK-64-NEXT:    br label %[[COND_END]]801// CK-64:       [[COND_END]]:802// CK-64-NEXT:    [[COND:%.*]] = phi i32 [ 9, %[[COND_TRUE]] ], [ [[TMP3]], %[[COND_FALSE]] ]803// CK-64-NEXT:    store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4804// CK-64-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4805// CK-64-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4806// CK-64-NEXT:    br label %[[OMP_INNER_FOR_COND:.*]]807// CK-64:       [[OMP_INNER_FOR_COND]]:808// CK-64-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4809// CK-64-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4810// CK-64-NEXT:    [[CMP1:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]811// CK-64-NEXT:    br i1 [[CMP1]], label %[[OMP_INNER_FOR_BODY:.*]], label %[[OMP_INNER_FOR_END:.*]]812// CK-64:       [[OMP_INNER_FOR_BODY]]:813// CK-64-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4814// CK-64-NEXT:    [[TMP8:%.*]] = zext i32 [[TMP7]] to i64815// CK-64-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4816// CK-64-NEXT:    [[TMP10:%.*]] = zext i32 [[TMP9]] to i64817// CK-64-NEXT:    [[TMP11:%.*]] = load i32, ptr [[X_ADDR]], align 4818// CK-64-NEXT:    store i32 [[TMP11]], ptr [[X_CASTED]], align 4819// CK-64-NEXT:    [[TMP12:%.*]] = load i64, ptr [[X_CASTED]], align 8820// CK-64-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB1]], i32 3, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.omp_outlined.omp_outlined, i64 [[TMP8]], i64 [[TMP10]], i64 [[TMP12]])821// CK-64-NEXT:    br label %[[OMP_INNER_FOR_INC:.*]]822// CK-64:       [[OMP_INNER_FOR_INC]]:823// CK-64-NEXT:    [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4824// CK-64-NEXT:    [[TMP14:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4825// CK-64-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP13]], [[TMP14]]826// CK-64-NEXT:    store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4827// CK-64-NEXT:    br label %[[OMP_INNER_FOR_COND]]828// CK-64:       [[OMP_INNER_FOR_END]]:829// CK-64-NEXT:    br label %[[OMP_LOOP_EXIT:.*]]830// CK-64:       [[OMP_LOOP_EXIT]]:831// CK-64-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP1]])832// CK-64-NEXT:    ret void833//834//835// CK-64-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.omp_outlined.omp_outlined(836// CK-64-SAME: ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i64 [[DOTPREVIOUS_LB_:%.*]], i64 [[DOTPREVIOUS_UB_:%.*]], i64 [[X:%.*]]) #[[ATTR1]] {837// CK-64-NEXT:  [[ENTRY:.*:]]838// CK-64-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8839// CK-64-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8840// CK-64-NEXT:    [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 8841// CK-64-NEXT:    [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 8842// CK-64-NEXT:    [[X_ADDR:%.*]] = alloca i64, align 8843// CK-64-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4844// CK-64-NEXT:    [[TMP:%.*]] = alloca i32, align 4845// CK-64-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4846// CK-64-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4847// CK-64-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4848// CK-64-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4849// CK-64-NEXT:    [[I:%.*]] = alloca i32, align 4850// CK-64-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8851// CK-64-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8852// CK-64-NEXT:    store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 8853// CK-64-NEXT:    store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 8854// CK-64-NEXT:    store i64 [[X]], ptr [[X_ADDR]], align 8855// CK-64-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4856// CK-64-NEXT:    store i32 9, ptr [[DOTOMP_UB]], align 4857// CK-64-NEXT:    [[TMP0:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 8858// CK-64-NEXT:    [[CONV:%.*]] = trunc i64 [[TMP0]] to i32859// CK-64-NEXT:    [[TMP1:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 8860// CK-64-NEXT:    [[CONV1:%.*]] = trunc i64 [[TMP1]] to i32861// CK-64-NEXT:    store i32 [[CONV]], ptr [[DOTOMP_LB]], align 4862// CK-64-NEXT:    store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 4863// CK-64-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4864// CK-64-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4865// CK-64-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8866// CK-64-NEXT:    [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 4867// CK-64-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB3:[0-9]+]], i32 [[TMP3]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)868// CK-64-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4869// CK-64-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 9870// CK-64-NEXT:    br i1 [[CMP]], label %[[COND_TRUE:.*]], label %[[COND_FALSE:.*]]871// CK-64:       [[COND_TRUE]]:872// CK-64-NEXT:    br label %[[COND_END:.*]]873// CK-64:       [[COND_FALSE]]:874// CK-64-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4875// CK-64-NEXT:    br label %[[COND_END]]876// CK-64:       [[COND_END]]:877// CK-64-NEXT:    [[COND:%.*]] = phi i32 [ 9, %[[COND_TRUE]] ], [ [[TMP5]], %[[COND_FALSE]] ]878// CK-64-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4879// CK-64-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4880// CK-64-NEXT:    store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 4881// CK-64-NEXT:    br label %[[OMP_INNER_FOR_COND:.*]]882// CK-64:       [[OMP_INNER_FOR_COND]]:883// CK-64-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4884// CK-64-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4885// CK-64-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]886// CK-64-NEXT:    br i1 [[CMP2]], label %[[OMP_INNER_FOR_BODY:.*]], label %[[OMP_INNER_FOR_END:.*]]887// CK-64:       [[OMP_INNER_FOR_BODY]]:888// CK-64-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4889// CK-64-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP9]], 1890// CK-64-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]891// CK-64-NEXT:    store i32 [[ADD]], ptr [[I]], align 4892// CK-64-NEXT:    [[TMP10:%.*]] = load i32, ptr [[X_ADDR]], align 4893// CK-64-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP10]], 1894// CK-64-NEXT:    store i32 [[ADD3]], ptr [[X_ADDR]], align 4895// CK-64-NEXT:    br label %[[OMP_BODY_CONTINUE:.*]]896// CK-64:       [[OMP_BODY_CONTINUE]]:897// CK-64-NEXT:    br label %[[OMP_INNER_FOR_INC:.*]]898// CK-64:       [[OMP_INNER_FOR_INC]]:899// CK-64-NEXT:    [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4900// CK-64-NEXT:    [[ADD4:%.*]] = add nsw i32 [[TMP11]], 1901// CK-64-NEXT:    store i32 [[ADD4]], ptr [[DOTOMP_IV]], align 4902// CK-64-NEXT:    br label %[[OMP_INNER_FOR_COND]]903// CK-64:       [[OMP_INNER_FOR_END]]:904// CK-64-NEXT:    br label %[[OMP_LOOP_EXIT:.*]]905// CK-64:       [[OMP_LOOP_EXIT]]:906// CK-64-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB3]], i32 [[TMP3]])907// CK-64-NEXT:    ret void908//909//910// CK-32-LABEL: define dso_local void @_Z4foo1i(911// CK-32-SAME: i32 [[A:%.*]]) #[[ATTR0:[0-9]+]] {912// CK-32-NEXT:  [[ENTRY:.*:]]913// CK-32-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4914// CK-32-NEXT:    [[D:%.*]] = alloca double, align 8915// CK-32-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 4916// CK-32-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 4917// CK-32-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 4918// CK-32-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8919// CK-32-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 4920// CK-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4921// CK-32-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double922// CK-32-NEXT:    store double [[CONV]], ptr [[D]], align 8923// CK-32-NEXT:    [[TMP1:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0924// CK-32-NEXT:    store ptr [[D]], ptr [[TMP1]], align 4925// CK-32-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0926// CK-32-NEXT:    store ptr [[D]], ptr [[TMP2]], align 4927// CK-32-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0928// CK-32-NEXT:    store ptr null, ptr [[TMP3]], align 4929// CK-32-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0930// CK-32-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0931// CK-32-NEXT:    [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0932// CK-32-NEXT:    store i32 3, ptr [[TMP6]], align 4933// CK-32-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1934// CK-32-NEXT:    store i32 1, ptr [[TMP7]], align 4935// CK-32-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2936// CK-32-NEXT:    store ptr [[TMP4]], ptr [[TMP8]], align 4937// CK-32-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3938// CK-32-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 4939// CK-32-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4940// CK-32-NEXT:    store ptr @.offload_sizes, ptr [[TMP10]], align 4941// CK-32-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5942// CK-32-NEXT:    store ptr @.offload_maptypes, ptr [[TMP11]], align 4943// CK-32-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6944// CK-32-NEXT:    store ptr null, ptr [[TMP12]], align 4945// CK-32-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7946// CK-32-NEXT:    store ptr null, ptr [[TMP13]], align 4947// CK-32-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8948// CK-32-NEXT:    store i64 0, ptr [[TMP14]], align 8949// CK-32-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9950// CK-32-NEXT:    store i64 0, ptr [[TMP15]], align 8951// CK-32-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10952// CK-32-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP16]], align 4953// CK-32-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11954// CK-32-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP17]], align 4955// CK-32-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12956// CK-32-NEXT:    store i32 0, ptr [[TMP18]], align 4957// CK-32-NEXT:    [[TMP19:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1:[0-9]+]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo1i_l23.region_id, ptr [[KERNEL_ARGS]])958// CK-32-NEXT:    [[TMP20:%.*]] = icmp ne i32 [[TMP19]], 0959// CK-32-NEXT:    br i1 [[TMP20]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]960// CK-32:       [[OMP_OFFLOAD_FAILED]]:961// CK-32-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo1i_l23(ptr [[D]]) #[[ATTR2:[0-9]+]]962// CK-32-NEXT:    br label %[[OMP_OFFLOAD_CONT]]963// CK-32:       [[OMP_OFFLOAD_CONT]]:964// CK-32-NEXT:    ret void965//966//967// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo1i_l23(968// CK-32-SAME: ptr nonnull align 4 dereferenceable(8) [[D:%.*]]) #[[ATTR1:[0-9]+]] {969// CK-32-NEXT:  [[ENTRY:.*:]]970// CK-32-NEXT:    [[D_ADDR:%.*]] = alloca ptr, align 4971// CK-32-NEXT:    [[D1:%.*]] = alloca double, align 8972// CK-32-NEXT:    store ptr [[D]], ptr [[D_ADDR]], align 4973// CK-32-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[D_ADDR]], align 4, !nonnull [[META20:![0-9]+]], !align [[META21:![0-9]+]]974// CK-32-NEXT:    [[TMP1:%.*]] = load double, ptr [[D1]], align 8975// CK-32-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+00976// CK-32-NEXT:    store double [[ADD]], ptr [[D1]], align 8977// CK-32-NEXT:    ret void978//979//980// CK-32-LABEL: define dso_local void @_Z4foo2v(981// CK-32-SAME: ) #[[ATTR0]] {982// CK-32-NEXT:  [[ENTRY:.*:]]983// CK-32-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 4984// CK-32-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 4985// CK-32-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 4986// CK-32-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 4987// CK-32-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8988// CK-32-NEXT:    [[TMP0:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0989// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[TMP0]], align 4990// CK-32-NEXT:    [[TMP1:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0991// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[TMP1]], align 4992// CK-32-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0993// CK-32-NEXT:    store ptr null, ptr [[TMP2]], align 4994// CK-32-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0995// CK-32-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0996// CK-32-NEXT:    [[TMP5:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0997// CK-32-NEXT:    store i32 3, ptr [[TMP5]], align 4998// CK-32-NEXT:    [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1999// CK-32-NEXT:    store i32 1, ptr [[TMP6]], align 41000// CK-32-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21001// CK-32-NEXT:    store ptr [[TMP3]], ptr [[TMP7]], align 41002// CK-32-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31003// CK-32-NEXT:    store ptr [[TMP4]], ptr [[TMP8]], align 41004// CK-32-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41005// CK-32-NEXT:    store ptr @.offload_sizes.1, ptr [[TMP9]], align 41006// CK-32-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51007// CK-32-NEXT:    store ptr @.offload_maptypes.2, ptr [[TMP10]], align 41008// CK-32-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61009// CK-32-NEXT:    store ptr null, ptr [[TMP11]], align 41010// CK-32-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71011// CK-32-NEXT:    store ptr null, ptr [[TMP12]], align 41012// CK-32-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81013// CK-32-NEXT:    store i64 0, ptr [[TMP13]], align 81014// CK-32-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91015// CK-32-NEXT:    store i64 0, ptr [[TMP14]], align 81016// CK-32-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101017// CK-32-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP15]], align 41018// CK-32-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111019// CK-32-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP16]], align 41020// CK-32-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121021// CK-32-NEXT:    store i32 0, ptr [[TMP17]], align 41022// CK-32-NEXT:    [[TMP18:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo2v_l32.region_id, ptr [[KERNEL_ARGS]])1023// CK-32-NEXT:    [[TMP19:%.*]] = icmp ne i32 [[TMP18]], 01024// CK-32-NEXT:    br i1 [[TMP19]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]1025// CK-32:       [[OMP_OFFLOAD_FAILED]]:1026// CK-32-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo2v_l32(ptr [[PVTARR]]) #[[ATTR2]]1027// CK-32-NEXT:    br label %[[OMP_OFFLOAD_CONT]]1028// CK-32:       [[OMP_OFFLOAD_CONT]]:1029// CK-32-NEXT:    ret void1030//1031//1032// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo2v_l32(1033// CK-32-SAME: ptr nonnull align 4 dereferenceable(40) [[PVTARR:%.*]]) #[[ATTR1]] {1034// CK-32-NEXT:  [[ENTRY:.*:]]1035// CK-32-NEXT:    [[PVTARR_ADDR:%.*]] = alloca ptr, align 41036// CK-32-NEXT:    [[PVTARR1:%.*]] = alloca [10 x i32], align 41037// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[PVTARR_ADDR]], align 41038// CK-32-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PVTARR_ADDR]], align 4, !nonnull [[META20]], !align [[META21]]1039// CK-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR1]], i32 0, i32 51040// CK-32-NEXT:    [[TMP1:%.*]] = load i32, ptr [[ARRAYIDX]], align 41041// CK-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP1]], 11042// CK-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41043// CK-32-NEXT:    ret void1044//1045//1046// CK-32-LABEL: define dso_local void @_Z4foo3v(1047// CK-32-SAME: ) #[[ATTR0]] {1048// CK-32-NEXT:  [[ENTRY:.*:]]1049// CK-32-NEXT:    [[PA:%.*]] = alloca ptr, align 41050// CK-32-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 41051// CK-32-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 41052// CK-32-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 41053// CK-32-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 81054// CK-32-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PA]], align 41055// CK-32-NEXT:    [[TMP1:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01056// CK-32-NEXT:    store ptr [[TMP0]], ptr [[TMP1]], align 41057// CK-32-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01058// CK-32-NEXT:    store ptr [[TMP0]], ptr [[TMP2]], align 41059// CK-32-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 01060// CK-32-NEXT:    store ptr null, ptr [[TMP3]], align 41061// CK-32-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01062// CK-32-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01063// CK-32-NEXT:    [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 01064// CK-32-NEXT:    store i32 3, ptr [[TMP6]], align 41065// CK-32-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11066// CK-32-NEXT:    store i32 1, ptr [[TMP7]], align 41067// CK-32-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21068// CK-32-NEXT:    store ptr [[TMP4]], ptr [[TMP8]], align 41069// CK-32-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31070// CK-32-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 41071// CK-32-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41072// CK-32-NEXT:    store ptr @.offload_sizes.3, ptr [[TMP10]], align 41073// CK-32-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51074// CK-32-NEXT:    store ptr @.offload_maptypes.4, ptr [[TMP11]], align 41075// CK-32-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61076// CK-32-NEXT:    store ptr null, ptr [[TMP12]], align 41077// CK-32-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71078// CK-32-NEXT:    store ptr null, ptr [[TMP13]], align 41079// CK-32-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81080// CK-32-NEXT:    store i64 0, ptr [[TMP14]], align 81081// CK-32-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91082// CK-32-NEXT:    store i64 0, ptr [[TMP15]], align 81083// CK-32-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101084// CK-32-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP16]], align 41085// CK-32-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111086// CK-32-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP17]], align 41087// CK-32-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121088// CK-32-NEXT:    store i32 0, ptr [[TMP18]], align 41089// CK-32-NEXT:    [[TMP19:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo3v_l41.region_id, ptr [[KERNEL_ARGS]])1090// CK-32-NEXT:    [[TMP20:%.*]] = icmp ne i32 [[TMP19]], 01091// CK-32-NEXT:    br i1 [[TMP20]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]1092// CK-32:       [[OMP_OFFLOAD_FAILED]]:1093// CK-32-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo3v_l41(ptr [[TMP0]]) #[[ATTR2]]1094// CK-32-NEXT:    br label %[[OMP_OFFLOAD_CONT]]1095// CK-32:       [[OMP_OFFLOAD_CONT]]:1096// CK-32-NEXT:    ret void1097//1098//1099// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo3v_l41(1100// CK-32-SAME: ptr [[PA:%.*]]) #[[ATTR1]] {1101// CK-32-NEXT:  [[ENTRY:.*:]]1102// CK-32-NEXT:    [[PA_ADDR:%.*]] = alloca ptr, align 41103// CK-32-NEXT:    [[PA1:%.*]] = alloca ptr, align 41104// CK-32-NEXT:    store ptr [[PA]], ptr [[PA_ADDR]], align 41105// CK-32-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PA1]], align 41106// CK-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds i32, ptr [[TMP0]], i32 501107// CK-32-NEXT:    [[TMP1:%.*]] = load i32, ptr [[ARRAYIDX]], align 41108// CK-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP1]], 11109// CK-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41110// CK-32-NEXT:    ret void1111//1112//1113// CK-32-LABEL: define dso_local void @_Z4foo4v(1114// CK-32-SAME: ) #[[ATTR0]] {1115// CK-32-NEXT:  [[ENTRY:.*:]]1116// CK-32-NEXT:    [[P:%.*]] = alloca i32, align 41117// CK-32-NEXT:    [[P_CASTED:%.*]] = alloca i32, align 41118// CK-32-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 41119// CK-32-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 41120// CK-32-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 41121// CK-32-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 81122// CK-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[P]], align 41123// CK-32-NEXT:    store i32 [[TMP0]], ptr [[P_CASTED]], align 41124// CK-32-NEXT:    [[TMP1:%.*]] = load i32, ptr [[P_CASTED]], align 41125// CK-32-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01126// CK-32-NEXT:    store i32 [[TMP1]], ptr [[TMP2]], align 41127// CK-32-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01128// CK-32-NEXT:    store i32 [[TMP1]], ptr [[TMP3]], align 41129// CK-32-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 01130// CK-32-NEXT:    store ptr null, ptr [[TMP4]], align 41131// CK-32-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01132// CK-32-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01133// CK-32-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 01134// CK-32-NEXT:    store i32 3, ptr [[TMP7]], align 41135// CK-32-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11136// CK-32-NEXT:    store i32 1, ptr [[TMP8]], align 41137// CK-32-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21138// CK-32-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 41139// CK-32-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31140// CK-32-NEXT:    store ptr [[TMP6]], ptr [[TMP10]], align 41141// CK-32-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41142// CK-32-NEXT:    store ptr @.offload_sizes.5, ptr [[TMP11]], align 41143// CK-32-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51144// CK-32-NEXT:    store ptr @.offload_maptypes.6, ptr [[TMP12]], align 41145// CK-32-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61146// CK-32-NEXT:    store ptr null, ptr [[TMP13]], align 41147// CK-32-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71148// CK-32-NEXT:    store ptr null, ptr [[TMP14]], align 41149// CK-32-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81150// CK-32-NEXT:    store i64 0, ptr [[TMP15]], align 81151// CK-32-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91152// CK-32-NEXT:    store i64 0, ptr [[TMP16]], align 81153// CK-32-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101154// CK-32-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP17]], align 41155// CK-32-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111156// CK-32-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP18]], align 41157// CK-32-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121158// CK-32-NEXT:    store i32 0, ptr [[TMP19]], align 41159// CK-32-NEXT:    [[TMP20:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo4v_l52.region_id, ptr [[KERNEL_ARGS]])1160// CK-32-NEXT:    [[TMP21:%.*]] = icmp ne i32 [[TMP20]], 01161// CK-32-NEXT:    br i1 [[TMP21]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]1162// CK-32:       [[OMP_OFFLOAD_FAILED]]:1163// CK-32-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo4v_l52(i32 [[TMP1]]) #[[ATTR2]]1164// CK-32-NEXT:    br label %[[OMP_OFFLOAD_CONT]]1165// CK-32:       [[OMP_OFFLOAD_CONT]]:1166// CK-32-NEXT:    ret void1167//1168//1169// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo4v_l52(1170// CK-32-SAME: i32 [[P:%.*]]) #[[ATTR1]] {1171// CK-32-NEXT:  [[ENTRY:.*:]]1172// CK-32-NEXT:    [[P_ADDR:%.*]] = alloca i32, align 41173// CK-32-NEXT:    store i32 [[P]], ptr [[P_ADDR]], align 41174// CK-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[P_ADDR]], align 41175// CK-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP0]], 11176// CK-32-NEXT:    store i32 [[INC]], ptr [[P_ADDR]], align 41177// CK-32-NEXT:    ret void1178//1179//1180// CK-32-LABEL: define dso_local void @_Z4foo5i(1181// CK-32-SAME: i32 [[A:%.*]]) #[[ATTR0]] {1182// CK-32-NEXT:  [[ENTRY:.*:]]1183// CK-32-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41184// CK-32-NEXT:    [[D:%.*]] = alloca double, align 81185// CK-32-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41186// CK-32-NEXT:    [[PA:%.*]] = alloca ptr, align 41187// CK-32-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x ptr], align 41188// CK-32-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x ptr], align 41189// CK-32-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x ptr], align 41190// CK-32-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 81191// CK-32-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41192// CK-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41193// CK-32-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1194// CK-32-NEXT:    store double [[CONV]], ptr [[D]], align 81195// CK-32-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[PA]], align 41196// CK-32-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01197// CK-32-NEXT:    store ptr [[D]], ptr [[TMP2]], align 41198// CK-32-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01199// CK-32-NEXT:    store ptr [[D]], ptr [[TMP3]], align 41200// CK-32-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 01201// CK-32-NEXT:    store ptr null, ptr [[TMP4]], align 41202// CK-32-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 11203// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[TMP5]], align 41204// CK-32-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 11205// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[TMP6]], align 41206// CK-32-NEXT:    [[TMP7:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 11207// CK-32-NEXT:    store ptr null, ptr [[TMP7]], align 41208// CK-32-NEXT:    [[TMP8:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 21209// CK-32-NEXT:    store ptr [[TMP1]], ptr [[TMP8]], align 41210// CK-32-NEXT:    [[TMP9:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 21211// CK-32-NEXT:    store ptr [[TMP1]], ptr [[TMP9]], align 41212// CK-32-NEXT:    [[TMP10:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 21213// CK-32-NEXT:    store ptr null, ptr [[TMP10]], align 41214// CK-32-NEXT:    [[TMP11:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01215// CK-32-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01216// CK-32-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 01217// CK-32-NEXT:    store i32 3, ptr [[TMP13]], align 41218// CK-32-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11219// CK-32-NEXT:    store i32 3, ptr [[TMP14]], align 41220// CK-32-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21221// CK-32-NEXT:    store ptr [[TMP11]], ptr [[TMP15]], align 41222// CK-32-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31223// CK-32-NEXT:    store ptr [[TMP12]], ptr [[TMP16]], align 41224// CK-32-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41225// CK-32-NEXT:    store ptr @.offload_sizes.7, ptr [[TMP17]], align 41226// CK-32-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51227// CK-32-NEXT:    store ptr @.offload_maptypes.8, ptr [[TMP18]], align 41228// CK-32-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61229// CK-32-NEXT:    store ptr null, ptr [[TMP19]], align 41230// CK-32-NEXT:    [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71231// CK-32-NEXT:    store ptr null, ptr [[TMP20]], align 41232// CK-32-NEXT:    [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81233// CK-32-NEXT:    store i64 0, ptr [[TMP21]], align 81234// CK-32-NEXT:    [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91235// CK-32-NEXT:    store i64 0, ptr [[TMP22]], align 81236// CK-32-NEXT:    [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101237// CK-32-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP23]], align 41238// CK-32-NEXT:    [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111239// CK-32-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP24]], align 41240// CK-32-NEXT:    [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121241// CK-32-NEXT:    store i32 0, ptr [[TMP25]], align 41242// CK-32-NEXT:    [[TMP26:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo5i_l66.region_id, ptr [[KERNEL_ARGS]])1243// CK-32-NEXT:    [[TMP27:%.*]] = icmp ne i32 [[TMP26]], 01244// CK-32-NEXT:    br i1 [[TMP27]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]1245// CK-32:       [[OMP_OFFLOAD_FAILED]]:1246// CK-32-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo5i_l66(ptr [[D]], ptr [[PVTARR]], ptr [[TMP1]]) #[[ATTR2]]1247// CK-32-NEXT:    br label %[[OMP_OFFLOAD_CONT]]1248// CK-32:       [[OMP_OFFLOAD_CONT]]:1249// CK-32-NEXT:    ret void1250//1251//1252// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo5i_l66(1253// CK-32-SAME: ptr nonnull align 4 dereferenceable(8) [[D:%.*]], ptr nonnull align 4 dereferenceable(40) [[PVTARR:%.*]], ptr [[PA:%.*]]) #[[ATTR1]] {1254// CK-32-NEXT:  [[ENTRY:.*:]]1255// CK-32-NEXT:    [[D_ADDR:%.*]] = alloca ptr, align 41256// CK-32-NEXT:    [[PVTARR_ADDR:%.*]] = alloca ptr, align 41257// CK-32-NEXT:    [[PA_ADDR:%.*]] = alloca ptr, align 41258// CK-32-NEXT:    [[D1:%.*]] = alloca double, align 81259// CK-32-NEXT:    [[PVTARR2:%.*]] = alloca [10 x i32], align 41260// CK-32-NEXT:    [[PA3:%.*]] = alloca ptr, align 41261// CK-32-NEXT:    store ptr [[D]], ptr [[D_ADDR]], align 41262// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[PVTARR_ADDR]], align 41263// CK-32-NEXT:    store ptr [[PA]], ptr [[PA_ADDR]], align 41264// CK-32-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[D_ADDR]], align 4, !nonnull [[META20]], !align [[META21]]1265// CK-32-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[PVTARR_ADDR]], align 4, !nonnull [[META20]], !align [[META21]]1266// CK-32-NEXT:    [[TMP2:%.*]] = load double, ptr [[D1]], align 81267// CK-32-NEXT:    [[ADD:%.*]] = fadd double [[TMP2]], 1.000000e+001268// CK-32-NEXT:    store double [[ADD]], ptr [[D1]], align 81269// CK-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR2]], i32 0, i32 51270// CK-32-NEXT:    [[TMP3:%.*]] = load i32, ptr [[ARRAYIDX]], align 41271// CK-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP3]], 11272// CK-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41273// CK-32-NEXT:    [[TMP4:%.*]] = load ptr, ptr [[PA3]], align 41274// CK-32-NEXT:    [[ARRAYIDX4:%.*]] = getelementptr inbounds i32, ptr [[TMP4]], i32 501275// CK-32-NEXT:    [[TMP5:%.*]] = load i32, ptr [[ARRAYIDX4]], align 41276// CK-32-NEXT:    [[INC5:%.*]] = add nsw i32 [[TMP5]], 11277// CK-32-NEXT:    store i32 [[INC5]], ptr [[ARRAYIDX4]], align 41278// CK-32-NEXT:    ret void1279//1280//1281// CK-32-LABEL: define dso_local void @_Z4foo6i(1282// CK-32-SAME: i32 [[A:%.*]]) #[[ATTR0]] {1283// CK-32-NEXT:  [[ENTRY:.*:]]1284// CK-32-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41285// CK-32-NEXT:    [[D:%.*]] = alloca double, align 81286// CK-32-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41287// CK-32-NEXT:    [[PA:%.*]] = alloca ptr, align 41288// CK-32-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x ptr], align 41289// CK-32-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x ptr], align 41290// CK-32-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x ptr], align 41291// CK-32-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 81292// CK-32-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41293// CK-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41294// CK-32-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1295// CK-32-NEXT:    store double [[CONV]], ptr [[D]], align 81296// CK-32-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[PA]], align 41297// CK-32-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01298// CK-32-NEXT:    store ptr [[D]], ptr [[TMP2]], align 41299// CK-32-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01300// CK-32-NEXT:    store ptr [[D]], ptr [[TMP3]], align 41301// CK-32-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 01302// CK-32-NEXT:    store ptr null, ptr [[TMP4]], align 41303// CK-32-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 11304// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[TMP5]], align 41305// CK-32-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 11306// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[TMP6]], align 41307// CK-32-NEXT:    [[TMP7:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 11308// CK-32-NEXT:    store ptr null, ptr [[TMP7]], align 41309// CK-32-NEXT:    [[TMP8:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 21310// CK-32-NEXT:    store ptr [[TMP1]], ptr [[TMP8]], align 41311// CK-32-NEXT:    [[TMP9:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 21312// CK-32-NEXT:    store ptr [[TMP1]], ptr [[TMP9]], align 41313// CK-32-NEXT:    [[TMP10:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 21314// CK-32-NEXT:    store ptr null, ptr [[TMP10]], align 41315// CK-32-NEXT:    [[TMP11:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01316// CK-32-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01317// CK-32-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 01318// CK-32-NEXT:    store i32 3, ptr [[TMP13]], align 41319// CK-32-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11320// CK-32-NEXT:    store i32 3, ptr [[TMP14]], align 41321// CK-32-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21322// CK-32-NEXT:    store ptr [[TMP11]], ptr [[TMP15]], align 41323// CK-32-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31324// CK-32-NEXT:    store ptr [[TMP12]], ptr [[TMP16]], align 41325// CK-32-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41326// CK-32-NEXT:    store ptr @.offload_sizes.9, ptr [[TMP17]], align 41327// CK-32-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51328// CK-32-NEXT:    store ptr @.offload_maptypes.10, ptr [[TMP18]], align 41329// CK-32-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61330// CK-32-NEXT:    store ptr null, ptr [[TMP19]], align 41331// CK-32-NEXT:    [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71332// CK-32-NEXT:    store ptr null, ptr [[TMP20]], align 41333// CK-32-NEXT:    [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81334// CK-32-NEXT:    store i64 0, ptr [[TMP21]], align 81335// CK-32-NEXT:    [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91336// CK-32-NEXT:    store i64 0, ptr [[TMP22]], align 81337// CK-32-NEXT:    [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101338// CK-32-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP23]], align 41339// CK-32-NEXT:    [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111340// CK-32-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP24]], align 41341// CK-32-NEXT:    [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121342// CK-32-NEXT:    store i32 0, ptr [[TMP25]], align 41343// CK-32-NEXT:    [[TMP26:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo6i_l82.region_id, ptr [[KERNEL_ARGS]])1344// CK-32-NEXT:    [[TMP27:%.*]] = icmp ne i32 [[TMP26]], 01345// CK-32-NEXT:    br i1 [[TMP27]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]1346// CK-32:       [[OMP_OFFLOAD_FAILED]]:1347// CK-32-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo6i_l82(ptr [[D]], ptr [[PVTARR]], ptr [[TMP1]]) #[[ATTR2]]1348// CK-32-NEXT:    br label %[[OMP_OFFLOAD_CONT]]1349// CK-32:       [[OMP_OFFLOAD_CONT]]:1350// CK-32-NEXT:    ret void1351//1352//1353// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo6i_l82(1354// CK-32-SAME: ptr nonnull align 4 dereferenceable(8) [[D:%.*]], ptr nonnull align 4 dereferenceable(40) [[PVTARR:%.*]], ptr [[PA:%.*]]) #[[ATTR1]] {1355// CK-32-NEXT:  [[ENTRY:.*:]]1356// CK-32-NEXT:    [[D_ADDR:%.*]] = alloca ptr, align 41357// CK-32-NEXT:    [[PVTARR_ADDR:%.*]] = alloca ptr, align 41358// CK-32-NEXT:    [[PA_ADDR:%.*]] = alloca ptr, align 41359// CK-32-NEXT:    [[D1:%.*]] = alloca double, align 81360// CK-32-NEXT:    store ptr [[D]], ptr [[D_ADDR]], align 41361// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[PVTARR_ADDR]], align 41362// CK-32-NEXT:    store ptr [[PA]], ptr [[PA_ADDR]], align 41363// CK-32-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[D_ADDR]], align 4, !nonnull [[META20]], !align [[META21]]1364// CK-32-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[PVTARR_ADDR]], align 4, !nonnull [[META20]], !align [[META21]]1365// CK-32-NEXT:    [[TMP2:%.*]] = load double, ptr [[TMP0]], align 81366// CK-32-NEXT:    store double [[TMP2]], ptr [[D1]], align 81367// CK-32-NEXT:    [[TMP3:%.*]] = load double, ptr [[D1]], align 81368// CK-32-NEXT:    [[ADD:%.*]] = fadd double [[TMP3]], 1.000000e+001369// CK-32-NEXT:    store double [[ADD]], ptr [[D1]], align 81370// CK-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[TMP1]], i32 0, i32 51371// CK-32-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX]], align 41372// CK-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP4]], 11373// CK-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41374// CK-32-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[PA_ADDR]], align 41375// CK-32-NEXT:    [[ARRAYIDX2:%.*]] = getelementptr inbounds i32, ptr [[TMP5]], i32 501376// CK-32-NEXT:    [[TMP6:%.*]] = load i32, ptr [[ARRAYIDX2]], align 41377// CK-32-NEXT:    [[INC3:%.*]] = add nsw i32 [[TMP6]], 11378// CK-32-NEXT:    store i32 [[INC3]], ptr [[ARRAYIDX2]], align 41379// CK-32-NEXT:    ret void1380//1381//1382// CK-32-LABEL: define dso_local void @_Z4foo7i(1383// CK-32-SAME: i32 [[A:%.*]]) #[[ATTR0]] {1384// CK-32-NEXT:  [[ENTRY:.*:]]1385// CK-32-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41386// CK-32-NEXT:    [[D:%.*]] = alloca double, align 81387// CK-32-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41388// CK-32-NEXT:    [[PA:%.*]] = alloca ptr, align 41389// CK-32-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x ptr], align 41390// CK-32-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x ptr], align 41391// CK-32-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x ptr], align 41392// CK-32-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 81393// CK-32-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41394// CK-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41395// CK-32-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1396// CK-32-NEXT:    store double [[CONV]], ptr [[D]], align 81397// CK-32-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[PA]], align 41398// CK-32-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01399// CK-32-NEXT:    store ptr [[D]], ptr [[TMP2]], align 41400// CK-32-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01401// CK-32-NEXT:    store ptr [[D]], ptr [[TMP3]], align 41402// CK-32-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 01403// CK-32-NEXT:    store ptr null, ptr [[TMP4]], align 41404// CK-32-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 11405// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[TMP5]], align 41406// CK-32-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 11407// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[TMP6]], align 41408// CK-32-NEXT:    [[TMP7:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 11409// CK-32-NEXT:    store ptr null, ptr [[TMP7]], align 41410// CK-32-NEXT:    [[TMP8:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 21411// CK-32-NEXT:    store ptr [[TMP1]], ptr [[TMP8]], align 41412// CK-32-NEXT:    [[TMP9:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 21413// CK-32-NEXT:    store ptr [[TMP1]], ptr [[TMP9]], align 41414// CK-32-NEXT:    [[TMP10:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 21415// CK-32-NEXT:    store ptr null, ptr [[TMP10]], align 41416// CK-32-NEXT:    [[TMP11:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01417// CK-32-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [3 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01418// CK-32-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 01419// CK-32-NEXT:    store i32 3, ptr [[TMP13]], align 41420// CK-32-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11421// CK-32-NEXT:    store i32 3, ptr [[TMP14]], align 41422// CK-32-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21423// CK-32-NEXT:    store ptr [[TMP11]], ptr [[TMP15]], align 41424// CK-32-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31425// CK-32-NEXT:    store ptr [[TMP12]], ptr [[TMP16]], align 41426// CK-32-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41427// CK-32-NEXT:    store ptr @.offload_sizes.11, ptr [[TMP17]], align 41428// CK-32-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51429// CK-32-NEXT:    store ptr @.offload_maptypes.12, ptr [[TMP18]], align 41430// CK-32-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61431// CK-32-NEXT:    store ptr null, ptr [[TMP19]], align 41432// CK-32-NEXT:    [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71433// CK-32-NEXT:    store ptr null, ptr [[TMP20]], align 41434// CK-32-NEXT:    [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81435// CK-32-NEXT:    store i64 0, ptr [[TMP21]], align 81436// CK-32-NEXT:    [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91437// CK-32-NEXT:    store i64 0, ptr [[TMP22]], align 81438// CK-32-NEXT:    [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101439// CK-32-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP23]], align 41440// CK-32-NEXT:    [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111441// CK-32-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP24]], align 41442// CK-32-NEXT:    [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121443// CK-32-NEXT:    store i32 0, ptr [[TMP25]], align 41444// CK-32-NEXT:    [[TMP26:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo7i_l98.region_id, ptr [[KERNEL_ARGS]])1445// CK-32-NEXT:    [[TMP27:%.*]] = icmp ne i32 [[TMP26]], 01446// CK-32-NEXT:    br i1 [[TMP27]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]1447// CK-32:       [[OMP_OFFLOAD_FAILED]]:1448// CK-32-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo7i_l98(ptr [[D]], ptr [[PVTARR]], ptr [[TMP1]]) #[[ATTR2]]1449// CK-32-NEXT:    br label %[[OMP_OFFLOAD_CONT]]1450// CK-32:       [[OMP_OFFLOAD_CONT]]:1451// CK-32-NEXT:    ret void1452//1453//1454// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo7i_l98(1455// CK-32-SAME: ptr nonnull align 4 dereferenceable(8) [[D:%.*]], ptr nonnull align 4 dereferenceable(40) [[PVTARR:%.*]], ptr [[PA:%.*]]) #[[ATTR1]] {1456// CK-32-NEXT:  [[ENTRY:.*:]]1457// CK-32-NEXT:    [[D_ADDR:%.*]] = alloca ptr, align 41458// CK-32-NEXT:    [[PVTARR_ADDR:%.*]] = alloca ptr, align 41459// CK-32-NEXT:    [[PA_ADDR:%.*]] = alloca ptr, align 41460// CK-32-NEXT:    [[D1:%.*]] = alloca double, align 81461// CK-32-NEXT:    [[PVTARR2:%.*]] = alloca [10 x i32], align 41462// CK-32-NEXT:    store ptr [[D]], ptr [[D_ADDR]], align 41463// CK-32-NEXT:    store ptr [[PVTARR]], ptr [[PVTARR_ADDR]], align 41464// CK-32-NEXT:    store ptr [[PA]], ptr [[PA_ADDR]], align 41465// CK-32-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[D_ADDR]], align 4, !nonnull [[META20]], !align [[META21]]1466// CK-32-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[PVTARR_ADDR]], align 4, !nonnull [[META20]], !align [[META21]]1467// CK-32-NEXT:    [[TMP2:%.*]] = load double, ptr [[TMP0]], align 81468// CK-32-NEXT:    store double [[TMP2]], ptr [[D1]], align 81469// CK-32-NEXT:    call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[PVTARR2]], ptr align 4 [[TMP1]], i32 40, i1 false)1470// CK-32-NEXT:    [[TMP3:%.*]] = load double, ptr [[D1]], align 81471// CK-32-NEXT:    [[ADD:%.*]] = fadd double [[TMP3]], 1.000000e+001472// CK-32-NEXT:    store double [[ADD]], ptr [[D1]], align 81473// CK-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR2]], i32 0, i32 51474// CK-32-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX]], align 41475// CK-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP4]], 11476// CK-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41477// CK-32-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[PA_ADDR]], align 41478// CK-32-NEXT:    [[ARRAYIDX3:%.*]] = getelementptr inbounds i32, ptr [[TMP5]], i32 501479// CK-32-NEXT:    [[TMP6:%.*]] = load i32, ptr [[ARRAYIDX3]], align 41480// CK-32-NEXT:    [[INC4:%.*]] = add nsw i32 [[TMP6]], 11481// CK-32-NEXT:    store i32 [[INC4]], ptr [[ARRAYIDX3]], align 41482// CK-32-NEXT:    ret void1483//1484//1485// CK-32-LABEL: define dso_local void @_Z4foo8v(1486// CK-32-SAME: ) #[[ATTR0]] {1487// CK-32-NEXT:  [[ENTRY:.*:]]1488// CK-32-NEXT:    [[X:%.*]] = alloca i32, align 41489// CK-32-NEXT:    [[X_CASTED:%.*]] = alloca i32, align 41490// CK-32-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 41491// CK-32-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 41492// CK-32-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 41493// CK-32-NEXT:    [[TMP:%.*]] = alloca i32, align 41494// CK-32-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 81495// CK-32-NEXT:    store i32 0, ptr [[X]], align 41496// CK-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[X]], align 41497// CK-32-NEXT:    store i32 [[TMP0]], ptr [[X_CASTED]], align 41498// CK-32-NEXT:    [[TMP1:%.*]] = load i32, ptr [[X_CASTED]], align 41499// CK-32-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01500// CK-32-NEXT:    store i32 [[TMP1]], ptr [[TMP2]], align 41501// CK-32-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01502// CK-32-NEXT:    store i32 [[TMP1]], ptr [[TMP3]], align 41503// CK-32-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 01504// CK-32-NEXT:    store ptr null, ptr [[TMP4]], align 41505// CK-32-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01506// CK-32-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01507// CK-32-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 01508// CK-32-NEXT:    store i32 3, ptr [[TMP7]], align 41509// CK-32-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11510// CK-32-NEXT:    store i32 1, ptr [[TMP8]], align 41511// CK-32-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21512// CK-32-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 41513// CK-32-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31514// CK-32-NEXT:    store ptr [[TMP6]], ptr [[TMP10]], align 41515// CK-32-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41516// CK-32-NEXT:    store ptr @.offload_sizes.13, ptr [[TMP11]], align 41517// CK-32-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51518// CK-32-NEXT:    store ptr @.offload_maptypes.14, ptr [[TMP12]], align 41519// CK-32-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61520// CK-32-NEXT:    store ptr null, ptr [[TMP13]], align 41521// CK-32-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71522// CK-32-NEXT:    store ptr null, ptr [[TMP14]], align 41523// CK-32-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81524// CK-32-NEXT:    store i64 10, ptr [[TMP15]], align 81525// CK-32-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91526// CK-32-NEXT:    store i64 0, ptr [[TMP16]], align 81527// CK-32-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101528// CK-32-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP17]], align 41529// CK-32-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111530// CK-32-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP18]], align 41531// CK-32-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121532// CK-32-NEXT:    store i32 0, ptr [[TMP19]], align 41533// CK-32-NEXT:    [[TMP20:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.region_id, ptr [[KERNEL_ARGS]])1534// CK-32-NEXT:    [[TMP21:%.*]] = icmp ne i32 [[TMP20]], 01535// CK-32-NEXT:    br i1 [[TMP21]], label %[[OMP_OFFLOAD_FAILED:.*]], label %[[OMP_OFFLOAD_CONT:.*]]1536// CK-32:       [[OMP_OFFLOAD_FAILED]]:1537// CK-32-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112(i32 [[TMP1]]) #[[ATTR2]]1538// CK-32-NEXT:    br label %[[OMP_OFFLOAD_CONT]]1539// CK-32:       [[OMP_OFFLOAD_CONT]]:1540// CK-32-NEXT:    ret void1541//1542//1543// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112(1544// CK-32-SAME: i32 [[X:%.*]]) #[[ATTR1]] {1545// CK-32-NEXT:  [[ENTRY:.*:]]1546// CK-32-NEXT:    [[X_ADDR:%.*]] = alloca i32, align 41547// CK-32-NEXT:    [[X_CASTED:%.*]] = alloca i32, align 41548// CK-32-NEXT:    store i32 [[X]], ptr [[X_ADDR]], align 41549// CK-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[X_ADDR]], align 41550// CK-32-NEXT:    store i32 [[TMP0]], ptr [[X_CASTED]], align 41551// CK-32-NEXT:    [[TMP1:%.*]] = load i32, ptr [[X_CASTED]], align 41552// CK-32-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.omp_outlined, i32 [[TMP1]])1553// CK-32-NEXT:    ret void1554//1555//1556// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.omp_outlined(1557// CK-32-SAME: ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i32 [[X:%.*]]) #[[ATTR1]] {1558// CK-32-NEXT:  [[ENTRY:.*:]]1559// CK-32-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 41560// CK-32-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 41561// CK-32-NEXT:    [[X_ADDR:%.*]] = alloca i32, align 41562// CK-32-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 41563// CK-32-NEXT:    [[TMP:%.*]] = alloca i32, align 41564// CK-32-NEXT:    [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 41565// CK-32-NEXT:    [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 41566// CK-32-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 41567// CK-32-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 41568// CK-32-NEXT:    [[I:%.*]] = alloca i32, align 41569// CK-32-NEXT:    [[X_CASTED:%.*]] = alloca i32, align 41570// CK-32-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 41571// CK-32-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 41572// CK-32-NEXT:    store i32 [[X]], ptr [[X_ADDR]], align 41573// CK-32-NEXT:    store i32 0, ptr [[DOTOMP_COMB_LB]], align 41574// CK-32-NEXT:    store i32 9, ptr [[DOTOMP_COMB_UB]], align 41575// CK-32-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 41576// CK-32-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 41577// CK-32-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 41578// CK-32-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 41579// CK-32-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB2:[0-9]+]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)1580// CK-32-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41581// CK-32-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 91582// CK-32-NEXT:    br i1 [[CMP]], label %[[COND_TRUE:.*]], label %[[COND_FALSE:.*]]1583// CK-32:       [[COND_TRUE]]:1584// CK-32-NEXT:    br label %[[COND_END:.*]]1585// CK-32:       [[COND_FALSE]]:1586// CK-32-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41587// CK-32-NEXT:    br label %[[COND_END]]1588// CK-32:       [[COND_END]]:1589// CK-32-NEXT:    [[COND:%.*]] = phi i32 [ 9, %[[COND_TRUE]] ], [ [[TMP3]], %[[COND_FALSE]] ]1590// CK-32-NEXT:    store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 41591// CK-32-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 41592// CK-32-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 41593// CK-32-NEXT:    br label %[[OMP_INNER_FOR_COND:.*]]1594// CK-32:       [[OMP_INNER_FOR_COND]]:1595// CK-32-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41596// CK-32-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41597// CK-32-NEXT:    [[CMP1:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]1598// CK-32-NEXT:    br i1 [[CMP1]], label %[[OMP_INNER_FOR_BODY:.*]], label %[[OMP_INNER_FOR_END:.*]]1599// CK-32:       [[OMP_INNER_FOR_BODY]]:1600// CK-32-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 41601// CK-32-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41602// CK-32-NEXT:    [[TMP9:%.*]] = load i32, ptr [[X_ADDR]], align 41603// CK-32-NEXT:    store i32 [[TMP9]], ptr [[X_CASTED]], align 41604// CK-32-NEXT:    [[TMP10:%.*]] = load i32, ptr [[X_CASTED]], align 41605// CK-32-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB1]], i32 3, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.omp_outlined.omp_outlined, i32 [[TMP7]], i32 [[TMP8]], i32 [[TMP10]])1606// CK-32-NEXT:    br label %[[OMP_INNER_FOR_INC:.*]]1607// CK-32:       [[OMP_INNER_FOR_INC]]:1608// CK-32-NEXT:    [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41609// CK-32-NEXT:    [[TMP12:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 41610// CK-32-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP11]], [[TMP12]]1611// CK-32-NEXT:    store i32 [[ADD]], ptr [[DOTOMP_IV]], align 41612// CK-32-NEXT:    br label %[[OMP_INNER_FOR_COND]]1613// CK-32:       [[OMP_INNER_FOR_END]]:1614// CK-32-NEXT:    br label %[[OMP_LOOP_EXIT:.*]]1615// CK-32:       [[OMP_LOOP_EXIT]]:1616// CK-32-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP1]])1617// CK-32-NEXT:    ret void1618//1619//1620// CK-32-LABEL: define internal void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z4foo8v_l112.omp_outlined.omp_outlined(1621// CK-32-SAME: ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i32 [[DOTPREVIOUS_LB_:%.*]], i32 [[DOTPREVIOUS_UB_:%.*]], i32 [[X:%.*]]) #[[ATTR1]] {1622// CK-32-NEXT:  [[ENTRY:.*:]]1623// CK-32-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 41624// CK-32-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 41625// CK-32-NEXT:    [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i32, align 41626// CK-32-NEXT:    [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i32, align 41627// CK-32-NEXT:    [[X_ADDR:%.*]] = alloca i32, align 41628// CK-32-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 41629// CK-32-NEXT:    [[TMP:%.*]] = alloca i32, align 41630// CK-32-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 41631// CK-32-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 41632// CK-32-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 41633// CK-32-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 41634// CK-32-NEXT:    [[I:%.*]] = alloca i32, align 41635// CK-32-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 41636// CK-32-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 41637// CK-32-NEXT:    store i32 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 41638// CK-32-NEXT:    store i32 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 41639// CK-32-NEXT:    store i32 [[X]], ptr [[X_ADDR]], align 41640// CK-32-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 41641// CK-32-NEXT:    store i32 9, ptr [[DOTOMP_UB]], align 41642// CK-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[DOTPREVIOUS_LB__ADDR]], align 41643// CK-32-NEXT:    [[TMP1:%.*]] = load i32, ptr [[DOTPREVIOUS_UB__ADDR]], align 41644// CK-32-NEXT:    store i32 [[TMP0]], ptr [[DOTOMP_LB]], align 41645// CK-32-NEXT:    store i32 [[TMP1]], ptr [[DOTOMP_UB]], align 41646// CK-32-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 41647// CK-32-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 41648// CK-32-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 41649// CK-32-NEXT:    [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 41650// CK-32-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB3:[0-9]+]], i32 [[TMP3]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)1651// CK-32-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41652// CK-32-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 91653// CK-32-NEXT:    br i1 [[CMP]], label %[[COND_TRUE:.*]], label %[[COND_FALSE:.*]]1654// CK-32:       [[COND_TRUE]]:1655// CK-32-NEXT:    br label %[[COND_END:.*]]1656// CK-32:       [[COND_FALSE]]:1657// CK-32-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41658// CK-32-NEXT:    br label %[[COND_END]]1659// CK-32:       [[COND_END]]:1660// CK-32-NEXT:    [[COND:%.*]] = phi i32 [ 9, %[[COND_TRUE]] ], [ [[TMP5]], %[[COND_FALSE]] ]1661// CK-32-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 41662// CK-32-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_LB]], align 41663// CK-32-NEXT:    store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 41664// CK-32-NEXT:    br label %[[OMP_INNER_FOR_COND:.*]]1665// CK-32:       [[OMP_INNER_FOR_COND]]:1666// CK-32-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41667// CK-32-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41668// CK-32-NEXT:    [[CMP1:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]1669// CK-32-NEXT:    br i1 [[CMP1]], label %[[OMP_INNER_FOR_BODY:.*]], label %[[OMP_INNER_FOR_END:.*]]1670// CK-32:       [[OMP_INNER_FOR_BODY]]:1671// CK-32-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41672// CK-32-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP9]], 11673// CK-32-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]1674// CK-32-NEXT:    store i32 [[ADD]], ptr [[I]], align 41675// CK-32-NEXT:    [[TMP10:%.*]] = load i32, ptr [[X_ADDR]], align 41676// CK-32-NEXT:    [[ADD2:%.*]] = add nsw i32 [[TMP10]], 11677// CK-32-NEXT:    store i32 [[ADD2]], ptr [[X_ADDR]], align 41678// CK-32-NEXT:    br label %[[OMP_BODY_CONTINUE:.*]]1679// CK-32:       [[OMP_BODY_CONTINUE]]:1680// CK-32-NEXT:    br label %[[OMP_INNER_FOR_INC:.*]]1681// CK-32:       [[OMP_INNER_FOR_INC]]:1682// CK-32-NEXT:    [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41683// CK-32-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP11]], 11684// CK-32-NEXT:    store i32 [[ADD3]], ptr [[DOTOMP_IV]], align 41685// CK-32-NEXT:    br label %[[OMP_INNER_FOR_COND]]1686// CK-32:       [[OMP_INNER_FOR_END]]:1687// CK-32-NEXT:    br label %[[OMP_LOOP_EXIT:.*]]1688// CK-32:       [[OMP_LOOP_EXIT]]:1689// CK-32-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB3]], i32 [[TMP3]])1690// CK-32-NEXT:    ret void1691//1692//1693// SIMD-ONLY-64-LABEL: define dso_local void @_Z4foo1i(1694// SIMD-ONLY-64-SAME: i32 signext [[A:%.*]]) #[[ATTR0:[0-9]+]] {1695// SIMD-ONLY-64-NEXT:  [[ENTRY:.*:]]1696// SIMD-ONLY-64-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41697// SIMD-ONLY-64-NEXT:    [[D:%.*]] = alloca double, align 81698// SIMD-ONLY-64-NEXT:    [[D1:%.*]] = alloca double, align 81699// SIMD-ONLY-64-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41700// SIMD-ONLY-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41701// SIMD-ONLY-64-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1702// SIMD-ONLY-64-NEXT:    store double [[CONV]], ptr [[D]], align 81703// SIMD-ONLY-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D1]], align 81704// SIMD-ONLY-64-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+001705// SIMD-ONLY-64-NEXT:    store double [[ADD]], ptr [[D1]], align 81706// SIMD-ONLY-64-NEXT:    ret void1707//1708//1709// SIMD-ONLY-64-LABEL: define dso_local void @_Z4foo2v(1710// SIMD-ONLY-64-SAME: ) #[[ATTR0]] {1711// SIMD-ONLY-64-NEXT:  [[ENTRY:.*:]]1712// SIMD-ONLY-64-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41713// SIMD-ONLY-64-NEXT:    [[PVTARR1:%.*]] = alloca [10 x i32], align 41714// SIMD-ONLY-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR1]], i64 0, i64 51715// SIMD-ONLY-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[ARRAYIDX]], align 41716// SIMD-ONLY-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP0]], 11717// SIMD-ONLY-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41718// SIMD-ONLY-64-NEXT:    ret void1719//1720//1721// SIMD-ONLY-64-LABEL: define dso_local void @_Z4foo3v(1722// SIMD-ONLY-64-SAME: ) #[[ATTR0]] {1723// SIMD-ONLY-64-NEXT:  [[ENTRY:.*:]]1724// SIMD-ONLY-64-NEXT:    [[PA:%.*]] = alloca ptr, align 81725// SIMD-ONLY-64-NEXT:    [[PA1:%.*]] = alloca ptr, align 81726// SIMD-ONLY-64-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PA1]], align 81727// SIMD-ONLY-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds i32, ptr [[TMP0]], i64 501728// SIMD-ONLY-64-NEXT:    [[TMP1:%.*]] = load i32, ptr [[ARRAYIDX]], align 41729// SIMD-ONLY-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP1]], 11730// SIMD-ONLY-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41731// SIMD-ONLY-64-NEXT:    ret void1732//1733//1734// SIMD-ONLY-64-LABEL: define dso_local void @_Z4foo4v(1735// SIMD-ONLY-64-SAME: ) #[[ATTR0]] {1736// SIMD-ONLY-64-NEXT:  [[ENTRY:.*:]]1737// SIMD-ONLY-64-NEXT:    [[P:%.*]] = alloca i32, align 41738// SIMD-ONLY-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[P]], align 41739// SIMD-ONLY-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP0]], 11740// SIMD-ONLY-64-NEXT:    store i32 [[INC]], ptr [[P]], align 41741// SIMD-ONLY-64-NEXT:    ret void1742//1743//1744// SIMD-ONLY-64-LABEL: define dso_local void @_Z4foo5i(1745// SIMD-ONLY-64-SAME: i32 signext [[A:%.*]]) #[[ATTR0]] {1746// SIMD-ONLY-64-NEXT:  [[ENTRY:.*:]]1747// SIMD-ONLY-64-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41748// SIMD-ONLY-64-NEXT:    [[D:%.*]] = alloca double, align 81749// SIMD-ONLY-64-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41750// SIMD-ONLY-64-NEXT:    [[PA:%.*]] = alloca ptr, align 81751// SIMD-ONLY-64-NEXT:    [[D1:%.*]] = alloca double, align 81752// SIMD-ONLY-64-NEXT:    [[PVTARR2:%.*]] = alloca [10 x i32], align 41753// SIMD-ONLY-64-NEXT:    [[PA3:%.*]] = alloca ptr, align 81754// SIMD-ONLY-64-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41755// SIMD-ONLY-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41756// SIMD-ONLY-64-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1757// SIMD-ONLY-64-NEXT:    store double [[CONV]], ptr [[D]], align 81758// SIMD-ONLY-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D1]], align 81759// SIMD-ONLY-64-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+001760// SIMD-ONLY-64-NEXT:    store double [[ADD]], ptr [[D1]], align 81761// SIMD-ONLY-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR2]], i64 0, i64 51762// SIMD-ONLY-64-NEXT:    [[TMP2:%.*]] = load i32, ptr [[ARRAYIDX]], align 41763// SIMD-ONLY-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 11764// SIMD-ONLY-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41765// SIMD-ONLY-64-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA3]], align 81766// SIMD-ONLY-64-NEXT:    [[ARRAYIDX4:%.*]] = getelementptr inbounds i32, ptr [[TMP3]], i64 501767// SIMD-ONLY-64-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX4]], align 41768// SIMD-ONLY-64-NEXT:    [[INC5:%.*]] = add nsw i32 [[TMP4]], 11769// SIMD-ONLY-64-NEXT:    store i32 [[INC5]], ptr [[ARRAYIDX4]], align 41770// SIMD-ONLY-64-NEXT:    ret void1771//1772//1773// SIMD-ONLY-64-LABEL: define dso_local void @_Z4foo6i(1774// SIMD-ONLY-64-SAME: i32 signext [[A:%.*]]) #[[ATTR0]] {1775// SIMD-ONLY-64-NEXT:  [[ENTRY:.*:]]1776// SIMD-ONLY-64-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41777// SIMD-ONLY-64-NEXT:    [[D:%.*]] = alloca double, align 81778// SIMD-ONLY-64-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41779// SIMD-ONLY-64-NEXT:    [[PA:%.*]] = alloca ptr, align 81780// SIMD-ONLY-64-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41781// SIMD-ONLY-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41782// SIMD-ONLY-64-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1783// SIMD-ONLY-64-NEXT:    store double [[CONV]], ptr [[D]], align 81784// SIMD-ONLY-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D]], align 81785// SIMD-ONLY-64-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+001786// SIMD-ONLY-64-NEXT:    store double [[ADD]], ptr [[D]], align 81787// SIMD-ONLY-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR]], i64 0, i64 51788// SIMD-ONLY-64-NEXT:    [[TMP2:%.*]] = load i32, ptr [[ARRAYIDX]], align 41789// SIMD-ONLY-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 11790// SIMD-ONLY-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41791// SIMD-ONLY-64-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA]], align 81792// SIMD-ONLY-64-NEXT:    [[ARRAYIDX1:%.*]] = getelementptr inbounds i32, ptr [[TMP3]], i64 501793// SIMD-ONLY-64-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX1]], align 41794// SIMD-ONLY-64-NEXT:    [[INC2:%.*]] = add nsw i32 [[TMP4]], 11795// SIMD-ONLY-64-NEXT:    store i32 [[INC2]], ptr [[ARRAYIDX1]], align 41796// SIMD-ONLY-64-NEXT:    ret void1797//1798//1799// SIMD-ONLY-64-LABEL: define dso_local void @_Z4foo7i(1800// SIMD-ONLY-64-SAME: i32 signext [[A:%.*]]) #[[ATTR0]] {1801// SIMD-ONLY-64-NEXT:  [[ENTRY:.*:]]1802// SIMD-ONLY-64-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41803// SIMD-ONLY-64-NEXT:    [[D:%.*]] = alloca double, align 81804// SIMD-ONLY-64-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41805// SIMD-ONLY-64-NEXT:    [[PA:%.*]] = alloca ptr, align 81806// SIMD-ONLY-64-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41807// SIMD-ONLY-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41808// SIMD-ONLY-64-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1809// SIMD-ONLY-64-NEXT:    store double [[CONV]], ptr [[D]], align 81810// SIMD-ONLY-64-NEXT:    [[TMP1:%.*]] = load double, ptr [[D]], align 81811// SIMD-ONLY-64-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+001812// SIMD-ONLY-64-NEXT:    store double [[ADD]], ptr [[D]], align 81813// SIMD-ONLY-64-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR]], i64 0, i64 51814// SIMD-ONLY-64-NEXT:    [[TMP2:%.*]] = load i32, ptr [[ARRAYIDX]], align 41815// SIMD-ONLY-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 11816// SIMD-ONLY-64-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41817// SIMD-ONLY-64-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA]], align 81818// SIMD-ONLY-64-NEXT:    [[ARRAYIDX1:%.*]] = getelementptr inbounds i32, ptr [[TMP3]], i64 501819// SIMD-ONLY-64-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX1]], align 41820// SIMD-ONLY-64-NEXT:    [[INC2:%.*]] = add nsw i32 [[TMP4]], 11821// SIMD-ONLY-64-NEXT:    store i32 [[INC2]], ptr [[ARRAYIDX1]], align 41822// SIMD-ONLY-64-NEXT:    ret void1823//1824//1825// SIMD-ONLY-64-LABEL: define dso_local void @_Z4foo8v(1826// SIMD-ONLY-64-SAME: ) #[[ATTR0]] {1827// SIMD-ONLY-64-NEXT:  [[ENTRY:.*:]]1828// SIMD-ONLY-64-NEXT:    [[X:%.*]] = alloca i32, align 41829// SIMD-ONLY-64-NEXT:    [[I:%.*]] = alloca i32, align 41830// SIMD-ONLY-64-NEXT:    store i32 0, ptr [[X]], align 41831// SIMD-ONLY-64-NEXT:    store i32 0, ptr [[I]], align 41832// SIMD-ONLY-64-NEXT:    br label %[[FOR_COND:.*]]1833// SIMD-ONLY-64:       [[FOR_COND]]:1834// SIMD-ONLY-64-NEXT:    [[TMP0:%.*]] = load i32, ptr [[I]], align 41835// SIMD-ONLY-64-NEXT:    [[CMP:%.*]] = icmp slt i32 [[TMP0]], 101836// SIMD-ONLY-64-NEXT:    br i1 [[CMP]], label %[[FOR_BODY:.*]], label %[[FOR_END:.*]]1837// SIMD-ONLY-64:       [[FOR_BODY]]:1838// SIMD-ONLY-64-NEXT:    [[TMP1:%.*]] = load i32, ptr [[X]], align 41839// SIMD-ONLY-64-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP1]], 11840// SIMD-ONLY-64-NEXT:    store i32 [[ADD]], ptr [[X]], align 41841// SIMD-ONLY-64-NEXT:    br label %[[FOR_INC:.*]]1842// SIMD-ONLY-64:       [[FOR_INC]]:1843// SIMD-ONLY-64-NEXT:    [[TMP2:%.*]] = load i32, ptr [[I]], align 41844// SIMD-ONLY-64-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 11845// SIMD-ONLY-64-NEXT:    store i32 [[INC]], ptr [[I]], align 41846// SIMD-ONLY-64-NEXT:    br label %[[FOR_COND]], !llvm.loop [[LOOP2:![0-9]+]]1847// SIMD-ONLY-64:       [[FOR_END]]:1848// SIMD-ONLY-64-NEXT:    ret void1849//1850//1851// SIMD-ONLY-32-LABEL: define dso_local void @_Z4foo1i(1852// SIMD-ONLY-32-SAME: i32 [[A:%.*]]) #[[ATTR0:[0-9]+]] {1853// SIMD-ONLY-32-NEXT:  [[ENTRY:.*:]]1854// SIMD-ONLY-32-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41855// SIMD-ONLY-32-NEXT:    [[D:%.*]] = alloca double, align 81856// SIMD-ONLY-32-NEXT:    [[D1:%.*]] = alloca double, align 81857// SIMD-ONLY-32-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41858// SIMD-ONLY-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41859// SIMD-ONLY-32-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1860// SIMD-ONLY-32-NEXT:    store double [[CONV]], ptr [[D]], align 81861// SIMD-ONLY-32-NEXT:    [[TMP1:%.*]] = load double, ptr [[D1]], align 81862// SIMD-ONLY-32-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+001863// SIMD-ONLY-32-NEXT:    store double [[ADD]], ptr [[D1]], align 81864// SIMD-ONLY-32-NEXT:    ret void1865//1866//1867// SIMD-ONLY-32-LABEL: define dso_local void @_Z4foo2v(1868// SIMD-ONLY-32-SAME: ) #[[ATTR0]] {1869// SIMD-ONLY-32-NEXT:  [[ENTRY:.*:]]1870// SIMD-ONLY-32-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41871// SIMD-ONLY-32-NEXT:    [[PVTARR1:%.*]] = alloca [10 x i32], align 41872// SIMD-ONLY-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR1]], i32 0, i32 51873// SIMD-ONLY-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[ARRAYIDX]], align 41874// SIMD-ONLY-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP0]], 11875// SIMD-ONLY-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41876// SIMD-ONLY-32-NEXT:    ret void1877//1878//1879// SIMD-ONLY-32-LABEL: define dso_local void @_Z4foo3v(1880// SIMD-ONLY-32-SAME: ) #[[ATTR0]] {1881// SIMD-ONLY-32-NEXT:  [[ENTRY:.*:]]1882// SIMD-ONLY-32-NEXT:    [[PA:%.*]] = alloca ptr, align 41883// SIMD-ONLY-32-NEXT:    [[PA1:%.*]] = alloca ptr, align 41884// SIMD-ONLY-32-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[PA1]], align 41885// SIMD-ONLY-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds i32, ptr [[TMP0]], i32 501886// SIMD-ONLY-32-NEXT:    [[TMP1:%.*]] = load i32, ptr [[ARRAYIDX]], align 41887// SIMD-ONLY-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP1]], 11888// SIMD-ONLY-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41889// SIMD-ONLY-32-NEXT:    ret void1890//1891//1892// SIMD-ONLY-32-LABEL: define dso_local void @_Z4foo4v(1893// SIMD-ONLY-32-SAME: ) #[[ATTR0]] {1894// SIMD-ONLY-32-NEXT:  [[ENTRY:.*:]]1895// SIMD-ONLY-32-NEXT:    [[P:%.*]] = alloca i32, align 41896// SIMD-ONLY-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[P]], align 41897// SIMD-ONLY-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP0]], 11898// SIMD-ONLY-32-NEXT:    store i32 [[INC]], ptr [[P]], align 41899// SIMD-ONLY-32-NEXT:    ret void1900//1901//1902// SIMD-ONLY-32-LABEL: define dso_local void @_Z4foo5i(1903// SIMD-ONLY-32-SAME: i32 [[A:%.*]]) #[[ATTR0]] {1904// SIMD-ONLY-32-NEXT:  [[ENTRY:.*:]]1905// SIMD-ONLY-32-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41906// SIMD-ONLY-32-NEXT:    [[D:%.*]] = alloca double, align 81907// SIMD-ONLY-32-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41908// SIMD-ONLY-32-NEXT:    [[PA:%.*]] = alloca ptr, align 41909// SIMD-ONLY-32-NEXT:    [[D1:%.*]] = alloca double, align 81910// SIMD-ONLY-32-NEXT:    [[PVTARR2:%.*]] = alloca [10 x i32], align 41911// SIMD-ONLY-32-NEXT:    [[PA3:%.*]] = alloca ptr, align 41912// SIMD-ONLY-32-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41913// SIMD-ONLY-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41914// SIMD-ONLY-32-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1915// SIMD-ONLY-32-NEXT:    store double [[CONV]], ptr [[D]], align 81916// SIMD-ONLY-32-NEXT:    [[TMP1:%.*]] = load double, ptr [[D1]], align 81917// SIMD-ONLY-32-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+001918// SIMD-ONLY-32-NEXT:    store double [[ADD]], ptr [[D1]], align 81919// SIMD-ONLY-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR2]], i32 0, i32 51920// SIMD-ONLY-32-NEXT:    [[TMP2:%.*]] = load i32, ptr [[ARRAYIDX]], align 41921// SIMD-ONLY-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 11922// SIMD-ONLY-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41923// SIMD-ONLY-32-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA3]], align 41924// SIMD-ONLY-32-NEXT:    [[ARRAYIDX4:%.*]] = getelementptr inbounds i32, ptr [[TMP3]], i32 501925// SIMD-ONLY-32-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX4]], align 41926// SIMD-ONLY-32-NEXT:    [[INC5:%.*]] = add nsw i32 [[TMP4]], 11927// SIMD-ONLY-32-NEXT:    store i32 [[INC5]], ptr [[ARRAYIDX4]], align 41928// SIMD-ONLY-32-NEXT:    ret void1929//1930//1931// SIMD-ONLY-32-LABEL: define dso_local void @_Z4foo6i(1932// SIMD-ONLY-32-SAME: i32 [[A:%.*]]) #[[ATTR0]] {1933// SIMD-ONLY-32-NEXT:  [[ENTRY:.*:]]1934// SIMD-ONLY-32-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41935// SIMD-ONLY-32-NEXT:    [[D:%.*]] = alloca double, align 81936// SIMD-ONLY-32-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41937// SIMD-ONLY-32-NEXT:    [[PA:%.*]] = alloca ptr, align 41938// SIMD-ONLY-32-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41939// SIMD-ONLY-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41940// SIMD-ONLY-32-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1941// SIMD-ONLY-32-NEXT:    store double [[CONV]], ptr [[D]], align 81942// SIMD-ONLY-32-NEXT:    [[TMP1:%.*]] = load double, ptr [[D]], align 81943// SIMD-ONLY-32-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+001944// SIMD-ONLY-32-NEXT:    store double [[ADD]], ptr [[D]], align 81945// SIMD-ONLY-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR]], i32 0, i32 51946// SIMD-ONLY-32-NEXT:    [[TMP2:%.*]] = load i32, ptr [[ARRAYIDX]], align 41947// SIMD-ONLY-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 11948// SIMD-ONLY-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41949// SIMD-ONLY-32-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA]], align 41950// SIMD-ONLY-32-NEXT:    [[ARRAYIDX1:%.*]] = getelementptr inbounds i32, ptr [[TMP3]], i32 501951// SIMD-ONLY-32-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX1]], align 41952// SIMD-ONLY-32-NEXT:    [[INC2:%.*]] = add nsw i32 [[TMP4]], 11953// SIMD-ONLY-32-NEXT:    store i32 [[INC2]], ptr [[ARRAYIDX1]], align 41954// SIMD-ONLY-32-NEXT:    ret void1955//1956//1957// SIMD-ONLY-32-LABEL: define dso_local void @_Z4foo7i(1958// SIMD-ONLY-32-SAME: i32 [[A:%.*]]) #[[ATTR0]] {1959// SIMD-ONLY-32-NEXT:  [[ENTRY:.*:]]1960// SIMD-ONLY-32-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 41961// SIMD-ONLY-32-NEXT:    [[D:%.*]] = alloca double, align 81962// SIMD-ONLY-32-NEXT:    [[PVTARR:%.*]] = alloca [10 x i32], align 41963// SIMD-ONLY-32-NEXT:    [[PA:%.*]] = alloca ptr, align 41964// SIMD-ONLY-32-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 41965// SIMD-ONLY-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41966// SIMD-ONLY-32-NEXT:    [[CONV:%.*]] = sitofp i32 [[TMP0]] to double1967// SIMD-ONLY-32-NEXT:    store double [[CONV]], ptr [[D]], align 81968// SIMD-ONLY-32-NEXT:    [[TMP1:%.*]] = load double, ptr [[D]], align 81969// SIMD-ONLY-32-NEXT:    [[ADD:%.*]] = fadd double [[TMP1]], 1.000000e+001970// SIMD-ONLY-32-NEXT:    store double [[ADD]], ptr [[D]], align 81971// SIMD-ONLY-32-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [10 x i32], ptr [[PVTARR]], i32 0, i32 51972// SIMD-ONLY-32-NEXT:    [[TMP2:%.*]] = load i32, ptr [[ARRAYIDX]], align 41973// SIMD-ONLY-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 11974// SIMD-ONLY-32-NEXT:    store i32 [[INC]], ptr [[ARRAYIDX]], align 41975// SIMD-ONLY-32-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[PA]], align 41976// SIMD-ONLY-32-NEXT:    [[ARRAYIDX1:%.*]] = getelementptr inbounds i32, ptr [[TMP3]], i32 501977// SIMD-ONLY-32-NEXT:    [[TMP4:%.*]] = load i32, ptr [[ARRAYIDX1]], align 41978// SIMD-ONLY-32-NEXT:    [[INC2:%.*]] = add nsw i32 [[TMP4]], 11979// SIMD-ONLY-32-NEXT:    store i32 [[INC2]], ptr [[ARRAYIDX1]], align 41980// SIMD-ONLY-32-NEXT:    ret void1981//1982//1983// SIMD-ONLY-32-LABEL: define dso_local void @_Z4foo8v(1984// SIMD-ONLY-32-SAME: ) #[[ATTR0]] {1985// SIMD-ONLY-32-NEXT:  [[ENTRY:.*:]]1986// SIMD-ONLY-32-NEXT:    [[X:%.*]] = alloca i32, align 41987// SIMD-ONLY-32-NEXT:    [[I:%.*]] = alloca i32, align 41988// SIMD-ONLY-32-NEXT:    store i32 0, ptr [[X]], align 41989// SIMD-ONLY-32-NEXT:    store i32 0, ptr [[I]], align 41990// SIMD-ONLY-32-NEXT:    br label %[[FOR_COND:.*]]1991// SIMD-ONLY-32:       [[FOR_COND]]:1992// SIMD-ONLY-32-NEXT:    [[TMP0:%.*]] = load i32, ptr [[I]], align 41993// SIMD-ONLY-32-NEXT:    [[CMP:%.*]] = icmp slt i32 [[TMP0]], 101994// SIMD-ONLY-32-NEXT:    br i1 [[CMP]], label %[[FOR_BODY:.*]], label %[[FOR_END:.*]]1995// SIMD-ONLY-32:       [[FOR_BODY]]:1996// SIMD-ONLY-32-NEXT:    [[TMP1:%.*]] = load i32, ptr [[X]], align 41997// SIMD-ONLY-32-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP1]], 11998// SIMD-ONLY-32-NEXT:    store i32 [[ADD]], ptr [[X]], align 41999// SIMD-ONLY-32-NEXT:    br label %[[FOR_INC:.*]]2000// SIMD-ONLY-32:       [[FOR_INC]]:2001// SIMD-ONLY-32-NEXT:    [[TMP2:%.*]] = load i32, ptr [[I]], align 42002// SIMD-ONLY-32-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP2]], 12003// SIMD-ONLY-32-NEXT:    store i32 [[INC]], ptr [[I]], align 42004// SIMD-ONLY-32-NEXT:    br label %[[FOR_COND]], !llvm.loop [[LOOP3:![0-9]+]]2005// SIMD-ONLY-32:       [[FOR_END]]:2006// SIMD-ONLY-32-NEXT:    ret void2007//2008//.2009// CK-64: [[META19]] = !{}2010// CK-64: [[META20]] = !{i64 4}2011//.2012// CK-32: [[META20]] = !{}2013// CK-32: [[META21]] = !{i64 4}2014//.2015// SIMD-ONLY-64: [[LOOP2]] = distinct !{[[LOOP2]], [[META3:![0-9]+]]}2016// SIMD-ONLY-64: [[META3]] = !{!"llvm.loop.mustprogress"}2017//.2018// SIMD-ONLY-32: [[LOOP3]] = distinct !{[[LOOP3]], [[META4:![0-9]+]]}2019// SIMD-ONLY-32: [[META4]] = !{!"llvm.loop.mustprogress"}2020//.2021