2769 lines · cpp
1// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs --replace-value-regex "__omp_offloading_[0-9a-z]+_[0-9a-z]+" "reduction_size[.].+[.]" "pl_cond[.].+[.|,]" --prefix-filecheck-ir-name _2// Test host codegen.3// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -verify -Wno-vla -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK14// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping5// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK16// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -verify -Wno-vla -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK37// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping8// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK39 10// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -verify -Wno-vla -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"11// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping12// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"13// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -verify -Wno-vla -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"14// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping15// RUN: %clang_cc1 -no-enable-noundef-analysis -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"16 17// RUN: %clang_cc1 -no-enable-noundef-analysis -verify -Wno-vla -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK918// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping19// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK920// RUN: %clang_cc1 -no-enable-noundef-analysis -verify -Wno-vla -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK1121// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping22// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK1123 24// RUN: %clang_cc1 -no-enable-noundef-analysis -verify -Wno-vla -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"25// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping26// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"27// RUN: %clang_cc1 -no-enable-noundef-analysis -verify -Wno-vla -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"28// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping29// RUN: %clang_cc1 -no-enable-noundef-analysis -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"30 31// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -verify -Wno-vla -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK1732// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping33// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK1734// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -verify -Wno-vla -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK1935// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping36// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK1937 38// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -verify -Wno-vla -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"39// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping40// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"41// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -verify -Wno-vla -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"42// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s -Wno-openmp-mapping43// RUN: %clang_cc1 -no-enable-noundef-analysis -DARRAY -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -include-pch %t -verify -Wno-vla %s -emit-llvm -o - -Wno-openmp-mapping | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"44// expected-no-diagnostics45#ifndef HEADER46#define HEADER47#ifndef ARRAY48struct St {49 int a, b;50 St() : a(0), b(0) {}51 St(const St &st) : a(st.a + st.b), b(0) {}52 ~St() {}53};54 55volatile int g __attribute__((aligned(128))) = 1212;56 57template <class T>58struct S {59 T f;60 S(T a) : f(a + g) {}61 S() : f(g) {}62 S(const S &s, St t = St()) : f(s.f + t.a) {}63 operator T() { return T(); }64 ~S() {}65};66 67 68template <typename T>69T tmain() {70 S<T> test;71 T t_var __attribute__((aligned(128))) = T();72 T vec[] __attribute__((aligned(128))) = {1, 2};73 S<T> s_arr[] __attribute__((aligned(128))) = {1, 2};74 S<T> var __attribute__((aligned(128))) (3);75 #pragma omp target76 #pragma omp teams firstprivate(t_var, vec, s_arr, var)77 {78 vec[0] = t_var;79 s_arr[0] = var;80 }81#pragma omp target82#pragma omp teams firstprivate(t_var)83 {}84 return T();85}86 87int main() {88 static int sivar;89#ifdef LAMBDA90 [&]() {91 #pragma omp target92 #pragma omp teams firstprivate(g, sivar)93 {94 g = 1;95 sivar = 2;96 [&]() {97 g = 2;98 sivar = 4;99 }();100 }101 }();102 return 0;103#else104 S<float> test;105 int t_var = 0;106 int vec[] = {1, 2};107 S<float> s_arr[] = {1, 2};108 S<float> var(3);109 #pragma omp target110 #pragma omp teams firstprivate(t_var, vec, s_arr, var, sivar)111 {112 vec[0] = t_var;113 s_arr[0] = var;114 sivar = 2;115 }116 #pragma omp target117 #pragma omp teams firstprivate(t_var)118 {}119 return tmain<int>();120#endif121}122 123 124 125 126 127 128 129 130 131 132 133 134 135#else136struct St {137 int a, b;138 St() : a(0), b(0) {}139 St(const St &) { }140 ~St() {}141 void St_func(St s[2], int n, long double vla1[n]) {142 double vla2[n][n] __attribute__((aligned(128)));143 a = b;144 #pragma omp target145 #pragma omp teams firstprivate(s, vla1, vla2)146 vla1[b] = vla2[1][n - 1] = a = b;147 }148};149 150void array_func(float a[3], St s[2], int n, long double vla1[n]) {151 double vla2[n][n] __attribute__((aligned(128)));152 #pragma omp target153 #pragma omp teams firstprivate(a, s, vla1, vla2)154 s[0].St_func(s, n, vla1);155 ;156}157 158#endif159#endif160// CHECK1-LABEL: define {{[^@]+}}@main161// CHECK1-SAME: () #[[ATTR0:[0-9]+]] {162// CHECK1-NEXT: entry:163// CHECK1-NEXT: [[RETVAL:%.*]] = alloca i32, align 4164// CHECK1-NEXT: [[REF_TMP:%.*]] = alloca [[CLASS_ANON:%.*]], align 1165// CHECK1-NEXT: store i32 0, ptr [[RETVAL]], align 4166// CHECK1-NEXT: call void @"_ZZ4mainENK3$_0clEv"(ptr nonnull align 1 dereferenceable(1) [[REF_TMP]])167// CHECK1-NEXT: ret i32 0168//169//170// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l91171// CHECK1-SAME: (i64 [[G:%.*]], i64 [[SIVAR:%.*]]) #[[ATTR2:[0-9]+]] {172// CHECK1-NEXT: entry:173// CHECK1-NEXT: [[G_ADDR:%.*]] = alloca i64, align 8174// CHECK1-NEXT: [[SIVAR_ADDR:%.*]] = alloca i64, align 8175// CHECK1-NEXT: [[G_CASTED:%.*]] = alloca i64, align 8176// CHECK1-NEXT: [[SIVAR_CASTED:%.*]] = alloca i64, align 8177// CHECK1-NEXT: store i64 [[G]], ptr [[G_ADDR]], align 8178// CHECK1-NEXT: store i64 [[SIVAR]], ptr [[SIVAR_ADDR]], align 8179// CHECK1-NEXT: [[TMP0:%.*]] = load i32, ptr [[G_ADDR]], align 4180// CHECK1-NEXT: store i32 [[TMP0]], ptr [[G_CASTED]], align 4181// CHECK1-NEXT: [[TMP1:%.*]] = load i64, ptr [[G_CASTED]], align 8182// CHECK1-NEXT: [[TMP2:%.*]] = load i32, ptr [[SIVAR_ADDR]], align 4183// CHECK1-NEXT: store i32 [[TMP2]], ptr [[SIVAR_CASTED]], align 4184// CHECK1-NEXT: [[TMP3:%.*]] = load i64, ptr [[SIVAR_CASTED]], align 8185// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1:[0-9]+]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l91.omp_outlined, i64 [[TMP1]], i64 [[TMP3]])186// CHECK1-NEXT: ret void187//188//189// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l91.omp_outlined190// CHECK1-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i64 [[G:%.*]], i64 [[SIVAR:%.*]]) #[[ATTR2]] {191// CHECK1-NEXT: entry:192// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8193// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8194// CHECK1-NEXT: [[G_ADDR:%.*]] = alloca i64, align 8195// CHECK1-NEXT: [[SIVAR_ADDR:%.*]] = alloca i64, align 8196// CHECK1-NEXT: [[REF_TMP:%.*]] = alloca [[CLASS_ANON_0:%.*]], align 8197// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8198// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8199// CHECK1-NEXT: store i64 [[G]], ptr [[G_ADDR]], align 8200// CHECK1-NEXT: store i64 [[SIVAR]], ptr [[SIVAR_ADDR]], align 8201// CHECK1-NEXT: store i32 1, ptr [[G_ADDR]], align 4202// CHECK1-NEXT: store i32 2, ptr [[SIVAR_ADDR]], align 4203// CHECK1-NEXT: [[TMP0:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 0204// CHECK1-NEXT: store ptr [[G_ADDR]], ptr [[TMP0]], align 8205// CHECK1-NEXT: [[TMP1:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 1206// CHECK1-NEXT: store ptr [[SIVAR_ADDR]], ptr [[TMP1]], align 8207// CHECK1-NEXT: call void @"_ZZZ4mainENK3$_0clEvENKUlvE_clEv"(ptr nonnull align 8 dereferenceable(16) [[REF_TMP]])208// CHECK1-NEXT: ret void209//210//211// CHECK3-LABEL: define {{[^@]+}}@main212// CHECK3-SAME: () #[[ATTR0:[0-9]+]] {213// CHECK3-NEXT: entry:214// CHECK3-NEXT: [[RETVAL:%.*]] = alloca i32, align 4215// CHECK3-NEXT: [[REF_TMP:%.*]] = alloca [[CLASS_ANON:%.*]], align 1216// CHECK3-NEXT: store i32 0, ptr [[RETVAL]], align 4217// CHECK3-NEXT: call void @"_ZZ4mainENK3$_0clEv"(ptr nonnull align 1 dereferenceable(1) [[REF_TMP]])218// CHECK3-NEXT: ret i32 0219//220//221// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l91222// CHECK3-SAME: (i32 [[G:%.*]], i32 [[SIVAR:%.*]]) #[[ATTR2:[0-9]+]] {223// CHECK3-NEXT: entry:224// CHECK3-NEXT: [[G_ADDR:%.*]] = alloca i32, align 4225// CHECK3-NEXT: [[SIVAR_ADDR:%.*]] = alloca i32, align 4226// CHECK3-NEXT: [[G_CASTED:%.*]] = alloca i32, align 4227// CHECK3-NEXT: [[SIVAR_CASTED:%.*]] = alloca i32, align 4228// CHECK3-NEXT: store i32 [[G]], ptr [[G_ADDR]], align 4229// CHECK3-NEXT: store i32 [[SIVAR]], ptr [[SIVAR_ADDR]], align 4230// CHECK3-NEXT: [[TMP0:%.*]] = load i32, ptr [[G_ADDR]], align 4231// CHECK3-NEXT: store i32 [[TMP0]], ptr [[G_CASTED]], align 4232// CHECK3-NEXT: [[TMP1:%.*]] = load i32, ptr [[G_CASTED]], align 4233// CHECK3-NEXT: [[TMP2:%.*]] = load i32, ptr [[SIVAR_ADDR]], align 4234// CHECK3-NEXT: store i32 [[TMP2]], ptr [[SIVAR_CASTED]], align 4235// CHECK3-NEXT: [[TMP3:%.*]] = load i32, ptr [[SIVAR_CASTED]], align 4236// CHECK3-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1:[0-9]+]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l91.omp_outlined, i32 [[TMP1]], i32 [[TMP3]])237// CHECK3-NEXT: ret void238//239//240// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l91.omp_outlined241// CHECK3-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i32 [[G:%.*]], i32 [[SIVAR:%.*]]) #[[ATTR2]] {242// CHECK3-NEXT: entry:243// CHECK3-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4244// CHECK3-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4245// CHECK3-NEXT: [[G_ADDR:%.*]] = alloca i32, align 4246// CHECK3-NEXT: [[SIVAR_ADDR:%.*]] = alloca i32, align 4247// CHECK3-NEXT: [[REF_TMP:%.*]] = alloca [[CLASS_ANON_0:%.*]], align 4248// CHECK3-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4249// CHECK3-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4250// CHECK3-NEXT: store i32 [[G]], ptr [[G_ADDR]], align 4251// CHECK3-NEXT: store i32 [[SIVAR]], ptr [[SIVAR_ADDR]], align 4252// CHECK3-NEXT: store i32 1, ptr [[G_ADDR]], align 4253// CHECK3-NEXT: store i32 2, ptr [[SIVAR_ADDR]], align 4254// CHECK3-NEXT: [[TMP0:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 0255// CHECK3-NEXT: store ptr [[G_ADDR]], ptr [[TMP0]], align 4256// CHECK3-NEXT: [[TMP1:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 1257// CHECK3-NEXT: store ptr [[SIVAR_ADDR]], ptr [[TMP1]], align 4258// CHECK3-NEXT: call void @"_ZZZ4mainENK3$_0clEvENKUlvE_clEv"(ptr nonnull align 4 dereferenceable(8) [[REF_TMP]])259// CHECK3-NEXT: ret void260//261//262// CHECK9-LABEL: define {{[^@]+}}@main263// CHECK9-SAME: () #[[ATTR0:[0-9]+]] {264// CHECK9-NEXT: entry:265// CHECK9-NEXT: [[RETVAL:%.*]] = alloca i32, align 4266// CHECK9-NEXT: [[TEST:%.*]] = alloca [[STRUCT_S:%.*]], align 4267// CHECK9-NEXT: [[T_VAR:%.*]] = alloca i32, align 4268// CHECK9-NEXT: [[VEC:%.*]] = alloca [2 x i32], align 4269// CHECK9-NEXT: [[S_ARR:%.*]] = alloca [2 x %struct.S], align 4270// CHECK9-NEXT: [[VAR:%.*]] = alloca [[STRUCT_S]], align 4271// CHECK9-NEXT: [[T_VAR_CASTED:%.*]] = alloca i64, align 8272// CHECK9-NEXT: [[SIVAR_CASTED:%.*]] = alloca i64, align 8273// CHECK9-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [5 x ptr], align 8274// CHECK9-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [5 x ptr], align 8275// CHECK9-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [5 x ptr], align 8276// CHECK9-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8277// CHECK9-NEXT: [[T_VAR_CASTED1:%.*]] = alloca i64, align 8278// CHECK9-NEXT: [[DOTOFFLOAD_BASEPTRS2:%.*]] = alloca [1 x ptr], align 8279// CHECK9-NEXT: [[DOTOFFLOAD_PTRS3:%.*]] = alloca [1 x ptr], align 8280// CHECK9-NEXT: [[DOTOFFLOAD_MAPPERS4:%.*]] = alloca [1 x ptr], align 8281// CHECK9-NEXT: [[KERNEL_ARGS5:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 8282// CHECK9-NEXT: store i32 0, ptr [[RETVAL]], align 4283// CHECK9-NEXT: call void @_ZN1SIfEC1Ev(ptr nonnull align 4 dereferenceable(4) [[TEST]])284// CHECK9-NEXT: store i32 0, ptr [[T_VAR]], align 4285// CHECK9-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 4 [[VEC]], ptr align 4 @__const.main.vec, i64 8, i1 false)286// CHECK9-NEXT: call void @_ZN1SIfEC1Ef(ptr nonnull align 4 dereferenceable(4) [[S_ARR]], float 1.000000e+00)287// CHECK9-NEXT: [[ARRAYINIT_ELEMENT:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[S_ARR]], i64 1288// CHECK9-NEXT: call void @_ZN1SIfEC1Ef(ptr nonnull align 4 dereferenceable(4) [[ARRAYINIT_ELEMENT]], float 2.000000e+00)289// CHECK9-NEXT: call void @_ZN1SIfEC1Ef(ptr nonnull align 4 dereferenceable(4) [[VAR]], float 3.000000e+00)290// CHECK9-NEXT: [[TMP0:%.*]] = load i32, ptr [[T_VAR]], align 4291// CHECK9-NEXT: store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 4292// CHECK9-NEXT: [[TMP1:%.*]] = load i64, ptr [[T_VAR_CASTED]], align 8293// CHECK9-NEXT: [[TMP2:%.*]] = load i32, ptr @_ZZ4mainE5sivar, align 4294// CHECK9-NEXT: store i32 [[TMP2]], ptr [[SIVAR_CASTED]], align 4295// CHECK9-NEXT: [[TMP3:%.*]] = load i64, ptr [[SIVAR_CASTED]], align 8296// CHECK9-NEXT: [[TMP4:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0297// CHECK9-NEXT: store i64 [[TMP1]], ptr [[TMP4]], align 8298// CHECK9-NEXT: [[TMP5:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0299// CHECK9-NEXT: store i64 [[TMP1]], ptr [[TMP5]], align 8300// CHECK9-NEXT: [[TMP6:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0301// CHECK9-NEXT: store ptr null, ptr [[TMP6]], align 8302// CHECK9-NEXT: [[TMP7:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1303// CHECK9-NEXT: store ptr [[VEC]], ptr [[TMP7]], align 8304// CHECK9-NEXT: [[TMP8:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1305// CHECK9-NEXT: store ptr [[VEC]], ptr [[TMP8]], align 8306// CHECK9-NEXT: [[TMP9:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1307// CHECK9-NEXT: store ptr null, ptr [[TMP9]], align 8308// CHECK9-NEXT: [[TMP10:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2309// CHECK9-NEXT: store ptr [[S_ARR]], ptr [[TMP10]], align 8310// CHECK9-NEXT: [[TMP11:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 2311// CHECK9-NEXT: store ptr [[S_ARR]], ptr [[TMP11]], align 8312// CHECK9-NEXT: [[TMP12:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2313// CHECK9-NEXT: store ptr null, ptr [[TMP12]], align 8314// CHECK9-NEXT: [[TMP13:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 3315// CHECK9-NEXT: store ptr [[VAR]], ptr [[TMP13]], align 8316// CHECK9-NEXT: [[TMP14:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 3317// CHECK9-NEXT: store ptr [[VAR]], ptr [[TMP14]], align 8318// CHECK9-NEXT: [[TMP15:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 3319// CHECK9-NEXT: store ptr null, ptr [[TMP15]], align 8320// CHECK9-NEXT: [[TMP16:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 4321// CHECK9-NEXT: store i64 [[TMP3]], ptr [[TMP16]], align 8322// CHECK9-NEXT: [[TMP17:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 4323// CHECK9-NEXT: store i64 [[TMP3]], ptr [[TMP17]], align 8324// CHECK9-NEXT: [[TMP18:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 4325// CHECK9-NEXT: store ptr null, ptr [[TMP18]], align 8326// CHECK9-NEXT: [[TMP19:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0327// CHECK9-NEXT: [[TMP20:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0328// CHECK9-NEXT: [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0329// CHECK9-NEXT: store i32 3, ptr [[TMP21]], align 4330// CHECK9-NEXT: [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1331// CHECK9-NEXT: store i32 5, ptr [[TMP22]], align 4332// CHECK9-NEXT: [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2333// CHECK9-NEXT: store ptr [[TMP19]], ptr [[TMP23]], align 8334// CHECK9-NEXT: [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3335// CHECK9-NEXT: store ptr [[TMP20]], ptr [[TMP24]], align 8336// CHECK9-NEXT: [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4337// CHECK9-NEXT: store ptr @.offload_sizes, ptr [[TMP25]], align 8338// CHECK9-NEXT: [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5339// CHECK9-NEXT: store ptr @.offload_maptypes, ptr [[TMP26]], align 8340// CHECK9-NEXT: [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6341// CHECK9-NEXT: store ptr null, ptr [[TMP27]], align 8342// CHECK9-NEXT: [[TMP28:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7343// CHECK9-NEXT: store ptr null, ptr [[TMP28]], align 8344// CHECK9-NEXT: [[TMP29:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8345// CHECK9-NEXT: store i64 0, ptr [[TMP29]], align 8346// CHECK9-NEXT: [[TMP30:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9347// CHECK9-NEXT: store i64 0, ptr [[TMP30]], align 8348// CHECK9-NEXT: [[TMP31:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10349// CHECK9-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP31]], align 4350// CHECK9-NEXT: [[TMP32:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11351// CHECK9-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP32]], align 4352// CHECK9-NEXT: [[TMP33:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12353// CHECK9-NEXT: store i32 0, ptr [[TMP33]], align 4354// CHECK9-NEXT: [[TMP34:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l109.region_id, ptr [[KERNEL_ARGS]])355// CHECK9-NEXT: [[TMP35:%.*]] = icmp ne i32 [[TMP34]], 0356// CHECK9-NEXT: br i1 [[TMP35]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]357// CHECK9: omp_offload.failed:358// CHECK9-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l109(i64 [[TMP1]], ptr [[VEC]], ptr [[S_ARR]], ptr [[VAR]], i64 [[TMP3]]) #[[ATTR4:[0-9]+]]359// CHECK9-NEXT: br label [[OMP_OFFLOAD_CONT]]360// CHECK9: omp_offload.cont:361// CHECK9-NEXT: [[TMP36:%.*]] = load i32, ptr [[T_VAR]], align 4362// CHECK9-NEXT: store i32 [[TMP36]], ptr [[T_VAR_CASTED1]], align 4363// CHECK9-NEXT: [[TMP37:%.*]] = load i64, ptr [[T_VAR_CASTED1]], align 8364// CHECK9-NEXT: [[TMP38:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS2]], i32 0, i32 0365// CHECK9-NEXT: store i64 [[TMP37]], ptr [[TMP38]], align 8366// CHECK9-NEXT: [[TMP39:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS3]], i32 0, i32 0367// CHECK9-NEXT: store i64 [[TMP37]], ptr [[TMP39]], align 8368// CHECK9-NEXT: [[TMP40:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS4]], i64 0, i64 0369// CHECK9-NEXT: store ptr null, ptr [[TMP40]], align 8370// CHECK9-NEXT: [[TMP41:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS2]], i32 0, i32 0371// CHECK9-NEXT: [[TMP42:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS3]], i32 0, i32 0372// CHECK9-NEXT: [[TMP43:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 0373// CHECK9-NEXT: store i32 3, ptr [[TMP43]], align 4374// CHECK9-NEXT: [[TMP44:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 1375// CHECK9-NEXT: store i32 1, ptr [[TMP44]], align 4376// CHECK9-NEXT: [[TMP45:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 2377// CHECK9-NEXT: store ptr [[TMP41]], ptr [[TMP45]], align 8378// CHECK9-NEXT: [[TMP46:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 3379// CHECK9-NEXT: store ptr [[TMP42]], ptr [[TMP46]], align 8380// CHECK9-NEXT: [[TMP47:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 4381// CHECK9-NEXT: store ptr @.offload_sizes.1, ptr [[TMP47]], align 8382// CHECK9-NEXT: [[TMP48:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 5383// CHECK9-NEXT: store ptr @.offload_maptypes.2, ptr [[TMP48]], align 8384// CHECK9-NEXT: [[TMP49:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 6385// CHECK9-NEXT: store ptr null, ptr [[TMP49]], align 8386// CHECK9-NEXT: [[TMP50:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 7387// CHECK9-NEXT: store ptr null, ptr [[TMP50]], align 8388// CHECK9-NEXT: [[TMP51:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 8389// CHECK9-NEXT: store i64 0, ptr [[TMP51]], align 8390// CHECK9-NEXT: [[TMP52:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 9391// CHECK9-NEXT: store i64 0, ptr [[TMP52]], align 8392// CHECK9-NEXT: [[TMP53:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 10393// CHECK9-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP53]], align 4394// CHECK9-NEXT: [[TMP54:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 11395// CHECK9-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP54]], align 4396// CHECK9-NEXT: [[TMP55:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 12397// CHECK9-NEXT: store i32 0, ptr [[TMP55]], align 4398// CHECK9-NEXT: [[TMP56:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l116.region_id, ptr [[KERNEL_ARGS5]])399// CHECK9-NEXT: [[TMP57:%.*]] = icmp ne i32 [[TMP56]], 0400// CHECK9-NEXT: br i1 [[TMP57]], label [[OMP_OFFLOAD_FAILED6:%.*]], label [[OMP_OFFLOAD_CONT7:%.*]]401// CHECK9: omp_offload.failed6:402// CHECK9-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l116(i64 [[TMP37]]) #[[ATTR4]]403// CHECK9-NEXT: br label [[OMP_OFFLOAD_CONT7]]404// CHECK9: omp_offload.cont7:405// CHECK9-NEXT: [[CALL:%.*]] = call signext i32 @_Z5tmainIiET_v()406// CHECK9-NEXT: store i32 [[CALL]], ptr [[RETVAL]], align 4407// CHECK9-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[VAR]]) #[[ATTR4]]408// CHECK9-NEXT: [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i32 0, i32 0409// CHECK9-NEXT: [[TMP58:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN]], i64 2410// CHECK9-NEXT: br label [[ARRAYDESTROY_BODY:%.*]]411// CHECK9: arraydestroy.body:412// CHECK9-NEXT: [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP58]], [[OMP_OFFLOAD_CONT7]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]413// CHECK9-NEXT: [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i64 -1414// CHECK9-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]415// CHECK9-NEXT: [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN]]416// CHECK9-NEXT: br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]417// CHECK9: arraydestroy.done8:418// CHECK9-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[TEST]]) #[[ATTR4]]419// CHECK9-NEXT: [[TMP59:%.*]] = load i32, ptr [[RETVAL]], align 4420// CHECK9-NEXT: ret i32 [[TMP59]]421//422//423// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC1Ev424// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1:[0-9]+]] comdat {425// CHECK9-NEXT: entry:426// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8427// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8428// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8429// CHECK9-NEXT: call void @_ZN1SIfEC2Ev(ptr nonnull align 4 dereferenceable(4) [[THIS1]])430// CHECK9-NEXT: ret void431//432//433// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC1Ef434// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], float [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat {435// CHECK9-NEXT: entry:436// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8437// CHECK9-NEXT: [[A_ADDR:%.*]] = alloca float, align 4438// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8439// CHECK9-NEXT: store float [[A]], ptr [[A_ADDR]], align 4440// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8441// CHECK9-NEXT: [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4442// CHECK9-NEXT: call void @_ZN1SIfEC2Ef(ptr nonnull align 4 dereferenceable(4) [[THIS1]], float [[TMP0]])443// CHECK9-NEXT: ret void444//445//446// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l109447// CHECK9-SAME: (i64 [[T_VAR:%.*]], ptr nonnull align 4 dereferenceable(8) [[VEC:%.*]], ptr nonnull align 4 dereferenceable(8) [[S_ARR:%.*]], ptr nonnull align 4 dereferenceable(4) [[VAR:%.*]], i64 [[SIVAR:%.*]]) #[[ATTR3:[0-9]+]] {448// CHECK9-NEXT: entry:449// CHECK9-NEXT: [[T_VAR_ADDR:%.*]] = alloca i64, align 8450// CHECK9-NEXT: [[VEC_ADDR:%.*]] = alloca ptr, align 8451// CHECK9-NEXT: [[S_ARR_ADDR:%.*]] = alloca ptr, align 8452// CHECK9-NEXT: [[VAR_ADDR:%.*]] = alloca ptr, align 8453// CHECK9-NEXT: [[SIVAR_ADDR:%.*]] = alloca i64, align 8454// CHECK9-NEXT: [[T_VAR_CASTED:%.*]] = alloca i64, align 8455// CHECK9-NEXT: [[SIVAR_CASTED:%.*]] = alloca i64, align 8456// CHECK9-NEXT: store i64 [[T_VAR]], ptr [[T_VAR_ADDR]], align 8457// CHECK9-NEXT: store ptr [[VEC]], ptr [[VEC_ADDR]], align 8458// CHECK9-NEXT: store ptr [[S_ARR]], ptr [[S_ARR_ADDR]], align 8459// CHECK9-NEXT: store ptr [[VAR]], ptr [[VAR_ADDR]], align 8460// CHECK9-NEXT: store i64 [[SIVAR]], ptr [[SIVAR_ADDR]], align 8461// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[VEC_ADDR]], align 8, !nonnull [[META11:![0-9]+]], !align [[META12:![0-9]+]]462// CHECK9-NEXT: [[TMP1:%.*]] = load ptr, ptr [[S_ARR_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]463// CHECK9-NEXT: [[TMP2:%.*]] = load ptr, ptr [[VAR_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]464// CHECK9-NEXT: [[TMP3:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 4465// CHECK9-NEXT: store i32 [[TMP3]], ptr [[T_VAR_CASTED]], align 4466// CHECK9-NEXT: [[TMP4:%.*]] = load i64, ptr [[T_VAR_CASTED]], align 8467// CHECK9-NEXT: [[TMP5:%.*]] = load i32, ptr [[SIVAR_ADDR]], align 4468// CHECK9-NEXT: store i32 [[TMP5]], ptr [[SIVAR_CASTED]], align 4469// CHECK9-NEXT: [[TMP6:%.*]] = load i64, ptr [[SIVAR_CASTED]], align 8470// CHECK9-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 5, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l109.omp_outlined, ptr [[TMP0]], i64 [[TMP4]], ptr [[TMP1]], ptr [[TMP2]], i64 [[TMP6]])471// CHECK9-NEXT: ret void472//473//474// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l109.omp_outlined475// CHECK9-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], ptr nonnull align 4 dereferenceable(8) [[VEC:%.*]], i64 [[T_VAR:%.*]], ptr nonnull align 4 dereferenceable(8) [[S_ARR:%.*]], ptr nonnull align 4 dereferenceable(4) [[VAR:%.*]], i64 [[SIVAR:%.*]]) #[[ATTR3]] {476// CHECK9-NEXT: entry:477// CHECK9-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8478// CHECK9-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8479// CHECK9-NEXT: [[VEC_ADDR:%.*]] = alloca ptr, align 8480// CHECK9-NEXT: [[T_VAR_ADDR:%.*]] = alloca i64, align 8481// CHECK9-NEXT: [[S_ARR_ADDR:%.*]] = alloca ptr, align 8482// CHECK9-NEXT: [[VAR_ADDR:%.*]] = alloca ptr, align 8483// CHECK9-NEXT: [[SIVAR_ADDR:%.*]] = alloca i64, align 8484// CHECK9-NEXT: [[VEC1:%.*]] = alloca [2 x i32], align 4485// CHECK9-NEXT: [[S_ARR2:%.*]] = alloca [2 x %struct.S], align 4486// CHECK9-NEXT: [[AGG_TMP:%.*]] = alloca [[STRUCT_ST:%.*]], align 4487// CHECK9-NEXT: [[VAR4:%.*]] = alloca [[STRUCT_S:%.*]], align 4488// CHECK9-NEXT: [[AGG_TMP5:%.*]] = alloca [[STRUCT_ST]], align 4489// CHECK9-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8490// CHECK9-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8491// CHECK9-NEXT: store ptr [[VEC]], ptr [[VEC_ADDR]], align 8492// CHECK9-NEXT: store i64 [[T_VAR]], ptr [[T_VAR_ADDR]], align 8493// CHECK9-NEXT: store ptr [[S_ARR]], ptr [[S_ARR_ADDR]], align 8494// CHECK9-NEXT: store ptr [[VAR]], ptr [[VAR_ADDR]], align 8495// CHECK9-NEXT: store i64 [[SIVAR]], ptr [[SIVAR_ADDR]], align 8496// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[VEC_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]497// CHECK9-NEXT: [[TMP1:%.*]] = load ptr, ptr [[S_ARR_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]498// CHECK9-NEXT: [[TMP2:%.*]] = load ptr, ptr [[VAR_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]499// CHECK9-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 4 [[VEC1]], ptr align 4 [[TMP0]], i64 8, i1 false)500// CHECK9-NEXT: [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR2]], i32 0, i32 0501// CHECK9-NEXT: [[TMP3:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN]], i64 2502// CHECK9-NEXT: [[OMP_ARRAYCPY_ISEMPTY:%.*]] = icmp eq ptr [[ARRAY_BEGIN]], [[TMP3]]503// CHECK9-NEXT: br i1 [[OMP_ARRAYCPY_ISEMPTY]], label [[OMP_ARRAYCPY_DONE3:%.*]], label [[OMP_ARRAYCPY_BODY:%.*]]504// CHECK9: omp.arraycpy.body:505// CHECK9-NEXT: [[OMP_ARRAYCPY_SRCELEMENTPAST:%.*]] = phi ptr [ [[TMP1]], [[ENTRY:%.*]] ], [ [[OMP_ARRAYCPY_SRC_ELEMENT:%.*]], [[OMP_ARRAYCPY_BODY]] ]506// CHECK9-NEXT: [[OMP_ARRAYCPY_DESTELEMENTPAST:%.*]] = phi ptr [ [[ARRAY_BEGIN]], [[ENTRY]] ], [ [[OMP_ARRAYCPY_DEST_ELEMENT:%.*]], [[OMP_ARRAYCPY_BODY]] ]507// CHECK9-NEXT: call void @_ZN2StC1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP]])508// CHECK9-NEXT: call void @_ZN1SIfEC1ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[OMP_ARRAYCPY_DESTELEMENTPAST]], ptr nonnull align 4 dereferenceable(4) [[OMP_ARRAYCPY_SRCELEMENTPAST]], ptr [[AGG_TMP]])509// CHECK9-NEXT: call void @_ZN2StD1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP]]) #[[ATTR4]]510// CHECK9-NEXT: [[OMP_ARRAYCPY_DEST_ELEMENT]] = getelementptr [[STRUCT_S]], ptr [[OMP_ARRAYCPY_DESTELEMENTPAST]], i32 1511// CHECK9-NEXT: [[OMP_ARRAYCPY_SRC_ELEMENT]] = getelementptr [[STRUCT_S]], ptr [[OMP_ARRAYCPY_SRCELEMENTPAST]], i32 1512// CHECK9-NEXT: [[OMP_ARRAYCPY_DONE:%.*]] = icmp eq ptr [[OMP_ARRAYCPY_DEST_ELEMENT]], [[TMP3]]513// CHECK9-NEXT: br i1 [[OMP_ARRAYCPY_DONE]], label [[OMP_ARRAYCPY_DONE3]], label [[OMP_ARRAYCPY_BODY]]514// CHECK9: omp.arraycpy.done3:515// CHECK9-NEXT: call void @_ZN2StC1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP5]])516// CHECK9-NEXT: call void @_ZN1SIfEC1ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[VAR4]], ptr nonnull align 4 dereferenceable(4) [[TMP2]], ptr [[AGG_TMP5]])517// CHECK9-NEXT: call void @_ZN2StD1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP5]]) #[[ATTR4]]518// CHECK9-NEXT: [[TMP4:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 4519// CHECK9-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[VEC1]], i64 0, i64 0520// CHECK9-NEXT: store i32 [[TMP4]], ptr [[ARRAYIDX]], align 4521// CHECK9-NEXT: [[ARRAYIDX6:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR2]], i64 0, i64 0522// CHECK9-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 4 [[ARRAYIDX6]], ptr align 4 [[VAR4]], i64 4, i1 false)523// CHECK9-NEXT: store i32 2, ptr [[SIVAR_ADDR]], align 4524// CHECK9-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[VAR4]]) #[[ATTR4]]525// CHECK9-NEXT: [[ARRAY_BEGIN7:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR2]], i32 0, i32 0526// CHECK9-NEXT: [[TMP5:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN7]], i64 2527// CHECK9-NEXT: br label [[ARRAYDESTROY_BODY:%.*]]528// CHECK9: arraydestroy.body:529// CHECK9-NEXT: [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP5]], [[OMP_ARRAYCPY_DONE3]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]530// CHECK9-NEXT: [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i64 -1531// CHECK9-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]532// CHECK9-NEXT: [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN7]]533// CHECK9-NEXT: br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]534// CHECK9: arraydestroy.done8:535// CHECK9-NEXT: ret void536//537//538// CHECK9-LABEL: define {{[^@]+}}@_ZN2StC1Ev539// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {540// CHECK9-NEXT: entry:541// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8542// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8543// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8544// CHECK9-NEXT: call void @_ZN2StC2Ev(ptr nonnull align 4 dereferenceable(8) [[THIS1]])545// CHECK9-NEXT: ret void546//547//548// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC1ERKS0_2St549// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], ptr nonnull align 4 dereferenceable(4) [[S:%.*]], ptr [[T:%.*]]) unnamed_addr #[[ATTR1]] comdat {550// CHECK9-NEXT: entry:551// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8552// CHECK9-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 8553// CHECK9-NEXT: [[T_INDIRECT_ADDR:%.*]] = alloca ptr, align 8554// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8555// CHECK9-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 8556// CHECK9-NEXT: store ptr [[T]], ptr [[T_INDIRECT_ADDR]], align 8557// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8558// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[S_ADDR]], align 8559// CHECK9-NEXT: call void @_ZN1SIfEC2ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[THIS1]], ptr nonnull align 4 dereferenceable(4) [[TMP0]], ptr [[T]])560// CHECK9-NEXT: ret void561//562//563// CHECK9-LABEL: define {{[^@]+}}@_ZN2StD1Ev564// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {565// CHECK9-NEXT: entry:566// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8567// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8568// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8569// CHECK9-NEXT: call void @_ZN2StD2Ev(ptr nonnull align 4 dereferenceable(8) [[THIS1]]) #[[ATTR4]]570// CHECK9-NEXT: ret void571//572//573// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfED1Ev574// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {575// CHECK9-NEXT: entry:576// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8577// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8578// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8579// CHECK9-NEXT: call void @_ZN1SIfED2Ev(ptr nonnull align 4 dereferenceable(4) [[THIS1]]) #[[ATTR4]]580// CHECK9-NEXT: ret void581//582//583// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l116584// CHECK9-SAME: (i64 [[T_VAR:%.*]]) #[[ATTR3]] {585// CHECK9-NEXT: entry:586// CHECK9-NEXT: [[T_VAR_ADDR:%.*]] = alloca i64, align 8587// CHECK9-NEXT: [[T_VAR_CASTED:%.*]] = alloca i64, align 8588// CHECK9-NEXT: store i64 [[T_VAR]], ptr [[T_VAR_ADDR]], align 8589// CHECK9-NEXT: [[TMP0:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 4590// CHECK9-NEXT: store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 4591// CHECK9-NEXT: [[TMP1:%.*]] = load i64, ptr [[T_VAR_CASTED]], align 8592// CHECK9-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l116.omp_outlined, i64 [[TMP1]])593// CHECK9-NEXT: ret void594//595//596// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l116.omp_outlined597// CHECK9-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i64 [[T_VAR:%.*]]) #[[ATTR3]] {598// CHECK9-NEXT: entry:599// CHECK9-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8600// CHECK9-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8601// CHECK9-NEXT: [[T_VAR_ADDR:%.*]] = alloca i64, align 8602// CHECK9-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8603// CHECK9-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8604// CHECK9-NEXT: store i64 [[T_VAR]], ptr [[T_VAR_ADDR]], align 8605// CHECK9-NEXT: ret void606//607//608// CHECK9-LABEL: define {{[^@]+}}@_Z5tmainIiET_v609// CHECK9-SAME: () #[[ATTR1]] comdat {610// CHECK9-NEXT: entry:611// CHECK9-NEXT: [[RETVAL:%.*]] = alloca i32, align 4612// CHECK9-NEXT: [[TEST:%.*]] = alloca [[STRUCT_S_0:%.*]], align 4613// CHECK9-NEXT: [[T_VAR:%.*]] = alloca i32, align 128614// CHECK9-NEXT: [[VEC:%.*]] = alloca [2 x i32], align 128615// CHECK9-NEXT: [[S_ARR:%.*]] = alloca [2 x %struct.S.0], align 128616// CHECK9-NEXT: [[VAR:%.*]] = alloca [[STRUCT_S_0]], align 128617// CHECK9-NEXT: [[T_VAR_CASTED:%.*]] = alloca i64, align 8618// CHECK9-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [4 x ptr], align 8619// CHECK9-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [4 x ptr], align 8620// CHECK9-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [4 x ptr], align 8621// CHECK9-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8622// CHECK9-NEXT: [[T_VAR_CASTED1:%.*]] = alloca i64, align 8623// CHECK9-NEXT: [[DOTOFFLOAD_BASEPTRS2:%.*]] = alloca [1 x ptr], align 8624// CHECK9-NEXT: [[DOTOFFLOAD_PTRS3:%.*]] = alloca [1 x ptr], align 8625// CHECK9-NEXT: [[DOTOFFLOAD_MAPPERS4:%.*]] = alloca [1 x ptr], align 8626// CHECK9-NEXT: [[KERNEL_ARGS5:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 8627// CHECK9-NEXT: call void @_ZN1SIiEC1Ev(ptr nonnull align 4 dereferenceable(4) [[TEST]])628// CHECK9-NEXT: store i32 0, ptr [[T_VAR]], align 128629// CHECK9-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 128 [[VEC]], ptr align 128 @__const._Z5tmainIiET_v.vec, i64 8, i1 false)630// CHECK9-NEXT: call void @_ZN1SIiEC1Ei(ptr nonnull align 4 dereferenceable(4) [[S_ARR]], i32 signext 1)631// CHECK9-NEXT: [[ARRAYINIT_ELEMENT:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[S_ARR]], i64 1632// CHECK9-NEXT: call void @_ZN1SIiEC1Ei(ptr nonnull align 4 dereferenceable(4) [[ARRAYINIT_ELEMENT]], i32 signext 2)633// CHECK9-NEXT: call void @_ZN1SIiEC1Ei(ptr nonnull align 4 dereferenceable(4) [[VAR]], i32 signext 3)634// CHECK9-NEXT: [[TMP0:%.*]] = load i32, ptr [[T_VAR]], align 128635// CHECK9-NEXT: store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 4636// CHECK9-NEXT: [[TMP1:%.*]] = load i64, ptr [[T_VAR_CASTED]], align 8637// CHECK9-NEXT: [[TMP2:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0638// CHECK9-NEXT: store i64 [[TMP1]], ptr [[TMP2]], align 8639// CHECK9-NEXT: [[TMP3:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0640// CHECK9-NEXT: store i64 [[TMP1]], ptr [[TMP3]], align 8641// CHECK9-NEXT: [[TMP4:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0642// CHECK9-NEXT: store ptr null, ptr [[TMP4]], align 8643// CHECK9-NEXT: [[TMP5:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1644// CHECK9-NEXT: store ptr [[VEC]], ptr [[TMP5]], align 8645// CHECK9-NEXT: [[TMP6:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1646// CHECK9-NEXT: store ptr [[VEC]], ptr [[TMP6]], align 8647// CHECK9-NEXT: [[TMP7:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1648// CHECK9-NEXT: store ptr null, ptr [[TMP7]], align 8649// CHECK9-NEXT: [[TMP8:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2650// CHECK9-NEXT: store ptr [[S_ARR]], ptr [[TMP8]], align 8651// CHECK9-NEXT: [[TMP9:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 2652// CHECK9-NEXT: store ptr [[S_ARR]], ptr [[TMP9]], align 8653// CHECK9-NEXT: [[TMP10:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2654// CHECK9-NEXT: store ptr null, ptr [[TMP10]], align 8655// CHECK9-NEXT: [[TMP11:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 3656// CHECK9-NEXT: store ptr [[VAR]], ptr [[TMP11]], align 8657// CHECK9-NEXT: [[TMP12:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 3658// CHECK9-NEXT: store ptr [[VAR]], ptr [[TMP12]], align 8659// CHECK9-NEXT: [[TMP13:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 3660// CHECK9-NEXT: store ptr null, ptr [[TMP13]], align 8661// CHECK9-NEXT: [[TMP14:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0662// CHECK9-NEXT: [[TMP15:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0663// CHECK9-NEXT: [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0664// CHECK9-NEXT: store i32 3, ptr [[TMP16]], align 4665// CHECK9-NEXT: [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1666// CHECK9-NEXT: store i32 4, ptr [[TMP17]], align 4667// CHECK9-NEXT: [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2668// CHECK9-NEXT: store ptr [[TMP14]], ptr [[TMP18]], align 8669// CHECK9-NEXT: [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3670// CHECK9-NEXT: store ptr [[TMP15]], ptr [[TMP19]], align 8671// CHECK9-NEXT: [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4672// CHECK9-NEXT: store ptr @.offload_sizes.3, ptr [[TMP20]], align 8673// CHECK9-NEXT: [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5674// CHECK9-NEXT: store ptr @.offload_maptypes.4, ptr [[TMP21]], align 8675// CHECK9-NEXT: [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6676// CHECK9-NEXT: store ptr null, ptr [[TMP22]], align 8677// CHECK9-NEXT: [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7678// CHECK9-NEXT: store ptr null, ptr [[TMP23]], align 8679// CHECK9-NEXT: [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8680// CHECK9-NEXT: store i64 0, ptr [[TMP24]], align 8681// CHECK9-NEXT: [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9682// CHECK9-NEXT: store i64 0, ptr [[TMP25]], align 8683// CHECK9-NEXT: [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10684// CHECK9-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP26]], align 4685// CHECK9-NEXT: [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11686// CHECK9-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP27]], align 4687// CHECK9-NEXT: [[TMP28:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12688// CHECK9-NEXT: store i32 0, ptr [[TMP28]], align 4689// CHECK9-NEXT: [[TMP29:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l75.region_id, ptr [[KERNEL_ARGS]])690// CHECK9-NEXT: [[TMP30:%.*]] = icmp ne i32 [[TMP29]], 0691// CHECK9-NEXT: br i1 [[TMP30]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]692// CHECK9: omp_offload.failed:693// CHECK9-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l75(i64 [[TMP1]], ptr [[VEC]], ptr [[S_ARR]], ptr [[VAR]]) #[[ATTR4]]694// CHECK9-NEXT: br label [[OMP_OFFLOAD_CONT]]695// CHECK9: omp_offload.cont:696// CHECK9-NEXT: [[TMP31:%.*]] = load i32, ptr [[T_VAR]], align 128697// CHECK9-NEXT: store i32 [[TMP31]], ptr [[T_VAR_CASTED1]], align 4698// CHECK9-NEXT: [[TMP32:%.*]] = load i64, ptr [[T_VAR_CASTED1]], align 8699// CHECK9-NEXT: [[TMP33:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS2]], i32 0, i32 0700// CHECK9-NEXT: store i64 [[TMP32]], ptr [[TMP33]], align 8701// CHECK9-NEXT: [[TMP34:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS3]], i32 0, i32 0702// CHECK9-NEXT: store i64 [[TMP32]], ptr [[TMP34]], align 8703// CHECK9-NEXT: [[TMP35:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS4]], i64 0, i64 0704// CHECK9-NEXT: store ptr null, ptr [[TMP35]], align 8705// CHECK9-NEXT: [[TMP36:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS2]], i32 0, i32 0706// CHECK9-NEXT: [[TMP37:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS3]], i32 0, i32 0707// CHECK9-NEXT: [[TMP38:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 0708// CHECK9-NEXT: store i32 3, ptr [[TMP38]], align 4709// CHECK9-NEXT: [[TMP39:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 1710// CHECK9-NEXT: store i32 1, ptr [[TMP39]], align 4711// CHECK9-NEXT: [[TMP40:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 2712// CHECK9-NEXT: store ptr [[TMP36]], ptr [[TMP40]], align 8713// CHECK9-NEXT: [[TMP41:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 3714// CHECK9-NEXT: store ptr [[TMP37]], ptr [[TMP41]], align 8715// CHECK9-NEXT: [[TMP42:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 4716// CHECK9-NEXT: store ptr @.offload_sizes.5, ptr [[TMP42]], align 8717// CHECK9-NEXT: [[TMP43:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 5718// CHECK9-NEXT: store ptr @.offload_maptypes.6, ptr [[TMP43]], align 8719// CHECK9-NEXT: [[TMP44:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 6720// CHECK9-NEXT: store ptr null, ptr [[TMP44]], align 8721// CHECK9-NEXT: [[TMP45:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 7722// CHECK9-NEXT: store ptr null, ptr [[TMP45]], align 8723// CHECK9-NEXT: [[TMP46:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 8724// CHECK9-NEXT: store i64 0, ptr [[TMP46]], align 8725// CHECK9-NEXT: [[TMP47:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 9726// CHECK9-NEXT: store i64 0, ptr [[TMP47]], align 8727// CHECK9-NEXT: [[TMP48:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 10728// CHECK9-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP48]], align 4729// CHECK9-NEXT: [[TMP49:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 11730// CHECK9-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP49]], align 4731// CHECK9-NEXT: [[TMP50:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 12732// CHECK9-NEXT: store i32 0, ptr [[TMP50]], align 4733// CHECK9-NEXT: [[TMP51:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l81.region_id, ptr [[KERNEL_ARGS5]])734// CHECK9-NEXT: [[TMP52:%.*]] = icmp ne i32 [[TMP51]], 0735// CHECK9-NEXT: br i1 [[TMP52]], label [[OMP_OFFLOAD_FAILED6:%.*]], label [[OMP_OFFLOAD_CONT7:%.*]]736// CHECK9: omp_offload.failed6:737// CHECK9-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l81(i64 [[TMP32]]) #[[ATTR4]]738// CHECK9-NEXT: br label [[OMP_OFFLOAD_CONT7]]739// CHECK9: omp_offload.cont7:740// CHECK9-NEXT: store i32 0, ptr [[RETVAL]], align 4741// CHECK9-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[VAR]]) #[[ATTR4]]742// CHECK9-NEXT: [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i32 0, i32 0743// CHECK9-NEXT: [[TMP53:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN]], i64 2744// CHECK9-NEXT: br label [[ARRAYDESTROY_BODY:%.*]]745// CHECK9: arraydestroy.body:746// CHECK9-NEXT: [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP53]], [[OMP_OFFLOAD_CONT7]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]747// CHECK9-NEXT: [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i64 -1748// CHECK9-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]749// CHECK9-NEXT: [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN]]750// CHECK9-NEXT: br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]751// CHECK9: arraydestroy.done8:752// CHECK9-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[TEST]]) #[[ATTR4]]753// CHECK9-NEXT: [[TMP54:%.*]] = load i32, ptr [[RETVAL]], align 4754// CHECK9-NEXT: ret i32 [[TMP54]]755//756//757// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC2Ev758// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {759// CHECK9-NEXT: entry:760// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8761// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8762// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8763// CHECK9-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0764// CHECK9-NEXT: [[TMP0:%.*]] = load volatile i32, ptr @g, align 128765// CHECK9-NEXT: [[CONV:%.*]] = sitofp i32 [[TMP0]] to float766// CHECK9-NEXT: store float [[CONV]], ptr [[F]], align 4767// CHECK9-NEXT: ret void768//769//770// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC2Ef771// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], float [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat {772// CHECK9-NEXT: entry:773// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8774// CHECK9-NEXT: [[A_ADDR:%.*]] = alloca float, align 4775// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8776// CHECK9-NEXT: store float [[A]], ptr [[A_ADDR]], align 4777// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8778// CHECK9-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0779// CHECK9-NEXT: [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4780// CHECK9-NEXT: [[TMP1:%.*]] = load volatile i32, ptr @g, align 128781// CHECK9-NEXT: [[CONV:%.*]] = sitofp i32 [[TMP1]] to float782// CHECK9-NEXT: [[ADD:%.*]] = fadd float [[TMP0]], [[CONV]]783// CHECK9-NEXT: store float [[ADD]], ptr [[F]], align 4784// CHECK9-NEXT: ret void785//786//787// CHECK9-LABEL: define {{[^@]+}}@_ZN2StC2Ev788// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {789// CHECK9-NEXT: entry:790// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8791// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8792// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8793// CHECK9-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[THIS1]], i32 0, i32 0794// CHECK9-NEXT: store i32 0, ptr [[A]], align 4795// CHECK9-NEXT: [[B:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[THIS1]], i32 0, i32 1796// CHECK9-NEXT: store i32 0, ptr [[B]], align 4797// CHECK9-NEXT: ret void798//799//800// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC2ERKS0_2St801// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], ptr nonnull align 4 dereferenceable(4) [[S:%.*]], ptr [[T:%.*]]) unnamed_addr #[[ATTR1]] comdat {802// CHECK9-NEXT: entry:803// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8804// CHECK9-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 8805// CHECK9-NEXT: [[T_INDIRECT_ADDR:%.*]] = alloca ptr, align 8806// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8807// CHECK9-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 8808// CHECK9-NEXT: store ptr [[T]], ptr [[T_INDIRECT_ADDR]], align 8809// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8810// CHECK9-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0811// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[S_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]812// CHECK9-NEXT: [[F2:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[TMP0]], i32 0, i32 0813// CHECK9-NEXT: [[TMP1:%.*]] = load float, ptr [[F2]], align 4814// CHECK9-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[T]], i32 0, i32 0815// CHECK9-NEXT: [[TMP2:%.*]] = load i32, ptr [[A]], align 4816// CHECK9-NEXT: [[CONV:%.*]] = sitofp i32 [[TMP2]] to float817// CHECK9-NEXT: [[ADD:%.*]] = fadd float [[TMP1]], [[CONV]]818// CHECK9-NEXT: store float [[ADD]], ptr [[F]], align 4819// CHECK9-NEXT: ret void820//821//822// CHECK9-LABEL: define {{[^@]+}}@_ZN2StD2Ev823// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {824// CHECK9-NEXT: entry:825// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8826// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8827// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8828// CHECK9-NEXT: ret void829//830//831// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfED2Ev832// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {833// CHECK9-NEXT: entry:834// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8835// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8836// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8837// CHECK9-NEXT: ret void838//839//840// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC1Ev841// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {842// CHECK9-NEXT: entry:843// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8844// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8845// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8846// CHECK9-NEXT: call void @_ZN1SIiEC2Ev(ptr nonnull align 4 dereferenceable(4) [[THIS1]])847// CHECK9-NEXT: ret void848//849//850// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC1Ei851// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], i32 signext [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat {852// CHECK9-NEXT: entry:853// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8854// CHECK9-NEXT: [[A_ADDR:%.*]] = alloca i32, align 4855// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8856// CHECK9-NEXT: store i32 [[A]], ptr [[A_ADDR]], align 4857// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8858// CHECK9-NEXT: [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4859// CHECK9-NEXT: call void @_ZN1SIiEC2Ei(ptr nonnull align 4 dereferenceable(4) [[THIS1]], i32 signext [[TMP0]])860// CHECK9-NEXT: ret void861//862//863// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l75864// CHECK9-SAME: (i64 [[T_VAR:%.*]], ptr nonnull align 4 dereferenceable(8) [[VEC:%.*]], ptr nonnull align 4 dereferenceable(8) [[S_ARR:%.*]], ptr nonnull align 4 dereferenceable(4) [[VAR:%.*]]) #[[ATTR3]] {865// CHECK9-NEXT: entry:866// CHECK9-NEXT: [[T_VAR_ADDR:%.*]] = alloca i64, align 8867// CHECK9-NEXT: [[VEC_ADDR:%.*]] = alloca ptr, align 8868// CHECK9-NEXT: [[S_ARR_ADDR:%.*]] = alloca ptr, align 8869// CHECK9-NEXT: [[VAR_ADDR:%.*]] = alloca ptr, align 8870// CHECK9-NEXT: [[T_VAR_CASTED:%.*]] = alloca i64, align 8871// CHECK9-NEXT: store i64 [[T_VAR]], ptr [[T_VAR_ADDR]], align 8872// CHECK9-NEXT: store ptr [[VEC]], ptr [[VEC_ADDR]], align 8873// CHECK9-NEXT: store ptr [[S_ARR]], ptr [[S_ARR_ADDR]], align 8874// CHECK9-NEXT: store ptr [[VAR]], ptr [[VAR_ADDR]], align 8875// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[VEC_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]876// CHECK9-NEXT: [[TMP1:%.*]] = load ptr, ptr [[S_ARR_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]877// CHECK9-NEXT: [[TMP2:%.*]] = load ptr, ptr [[VAR_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]878// CHECK9-NEXT: [[TMP3:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 4879// CHECK9-NEXT: store i32 [[TMP3]], ptr [[T_VAR_CASTED]], align 4880// CHECK9-NEXT: [[TMP4:%.*]] = load i64, ptr [[T_VAR_CASTED]], align 8881// CHECK9-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 4, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l75.omp_outlined, ptr [[TMP0]], i64 [[TMP4]], ptr [[TMP1]], ptr [[TMP2]])882// CHECK9-NEXT: ret void883//884//885// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l75.omp_outlined886// CHECK9-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], ptr nonnull align 4 dereferenceable(8) [[VEC:%.*]], i64 [[T_VAR:%.*]], ptr nonnull align 4 dereferenceable(8) [[S_ARR:%.*]], ptr nonnull align 4 dereferenceable(4) [[VAR:%.*]]) #[[ATTR3]] {887// CHECK9-NEXT: entry:888// CHECK9-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8889// CHECK9-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8890// CHECK9-NEXT: [[VEC_ADDR:%.*]] = alloca ptr, align 8891// CHECK9-NEXT: [[T_VAR_ADDR:%.*]] = alloca i64, align 8892// CHECK9-NEXT: [[S_ARR_ADDR:%.*]] = alloca ptr, align 8893// CHECK9-NEXT: [[VAR_ADDR:%.*]] = alloca ptr, align 8894// CHECK9-NEXT: [[VEC1:%.*]] = alloca [2 x i32], align 128895// CHECK9-NEXT: [[S_ARR2:%.*]] = alloca [2 x %struct.S.0], align 128896// CHECK9-NEXT: [[AGG_TMP:%.*]] = alloca [[STRUCT_ST:%.*]], align 4897// CHECK9-NEXT: [[VAR4:%.*]] = alloca [[STRUCT_S_0:%.*]], align 128898// CHECK9-NEXT: [[AGG_TMP5:%.*]] = alloca [[STRUCT_ST]], align 4899// CHECK9-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8900// CHECK9-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8901// CHECK9-NEXT: store ptr [[VEC]], ptr [[VEC_ADDR]], align 8902// CHECK9-NEXT: store i64 [[T_VAR]], ptr [[T_VAR_ADDR]], align 8903// CHECK9-NEXT: store ptr [[S_ARR]], ptr [[S_ARR_ADDR]], align 8904// CHECK9-NEXT: store ptr [[VAR]], ptr [[VAR_ADDR]], align 8905// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[VEC_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]906// CHECK9-NEXT: [[TMP1:%.*]] = load ptr, ptr [[S_ARR_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]907// CHECK9-NEXT: [[TMP2:%.*]] = load ptr, ptr [[VAR_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]908// CHECK9-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 128 [[VEC1]], ptr align 128 [[TMP0]], i64 8, i1 false)909// CHECK9-NEXT: [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR2]], i32 0, i32 0910// CHECK9-NEXT: [[TMP3:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN]], i64 2911// CHECK9-NEXT: [[OMP_ARRAYCPY_ISEMPTY:%.*]] = icmp eq ptr [[ARRAY_BEGIN]], [[TMP3]]912// CHECK9-NEXT: br i1 [[OMP_ARRAYCPY_ISEMPTY]], label [[OMP_ARRAYCPY_DONE3:%.*]], label [[OMP_ARRAYCPY_BODY:%.*]]913// CHECK9: omp.arraycpy.body:914// CHECK9-NEXT: [[OMP_ARRAYCPY_SRCELEMENTPAST:%.*]] = phi ptr [ [[TMP1]], [[ENTRY:%.*]] ], [ [[OMP_ARRAYCPY_SRC_ELEMENT:%.*]], [[OMP_ARRAYCPY_BODY]] ]915// CHECK9-NEXT: [[OMP_ARRAYCPY_DESTELEMENTPAST:%.*]] = phi ptr [ [[ARRAY_BEGIN]], [[ENTRY]] ], [ [[OMP_ARRAYCPY_DEST_ELEMENT:%.*]], [[OMP_ARRAYCPY_BODY]] ]916// CHECK9-NEXT: call void @_ZN2StC1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP]])917// CHECK9-NEXT: call void @_ZN1SIiEC1ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[OMP_ARRAYCPY_DESTELEMENTPAST]], ptr nonnull align 4 dereferenceable(4) [[OMP_ARRAYCPY_SRCELEMENTPAST]], ptr [[AGG_TMP]])918// CHECK9-NEXT: call void @_ZN2StD1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP]]) #[[ATTR4]]919// CHECK9-NEXT: [[OMP_ARRAYCPY_DEST_ELEMENT]] = getelementptr [[STRUCT_S_0]], ptr [[OMP_ARRAYCPY_DESTELEMENTPAST]], i32 1920// CHECK9-NEXT: [[OMP_ARRAYCPY_SRC_ELEMENT]] = getelementptr [[STRUCT_S_0]], ptr [[OMP_ARRAYCPY_SRCELEMENTPAST]], i32 1921// CHECK9-NEXT: [[OMP_ARRAYCPY_DONE:%.*]] = icmp eq ptr [[OMP_ARRAYCPY_DEST_ELEMENT]], [[TMP3]]922// CHECK9-NEXT: br i1 [[OMP_ARRAYCPY_DONE]], label [[OMP_ARRAYCPY_DONE3]], label [[OMP_ARRAYCPY_BODY]]923// CHECK9: omp.arraycpy.done3:924// CHECK9-NEXT: call void @_ZN2StC1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP5]])925// CHECK9-NEXT: call void @_ZN1SIiEC1ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[VAR4]], ptr nonnull align 4 dereferenceable(4) [[TMP2]], ptr [[AGG_TMP5]])926// CHECK9-NEXT: call void @_ZN2StD1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP5]]) #[[ATTR4]]927// CHECK9-NEXT: [[TMP4:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 4928// CHECK9-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[VEC1]], i64 0, i64 0929// CHECK9-NEXT: store i32 [[TMP4]], ptr [[ARRAYIDX]], align 128930// CHECK9-NEXT: [[ARRAYIDX6:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR2]], i64 0, i64 0931// CHECK9-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 128 [[ARRAYIDX6]], ptr align 128 [[VAR4]], i64 4, i1 false)932// CHECK9-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[VAR4]]) #[[ATTR4]]933// CHECK9-NEXT: [[ARRAY_BEGIN7:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR2]], i32 0, i32 0934// CHECK9-NEXT: [[TMP5:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN7]], i64 2935// CHECK9-NEXT: br label [[ARRAYDESTROY_BODY:%.*]]936// CHECK9: arraydestroy.body:937// CHECK9-NEXT: [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP5]], [[OMP_ARRAYCPY_DONE3]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]938// CHECK9-NEXT: [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i64 -1939// CHECK9-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]940// CHECK9-NEXT: [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN7]]941// CHECK9-NEXT: br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]942// CHECK9: arraydestroy.done8:943// CHECK9-NEXT: ret void944//945//946// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC1ERKS0_2St947// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], ptr nonnull align 4 dereferenceable(4) [[S:%.*]], ptr [[T:%.*]]) unnamed_addr #[[ATTR1]] comdat {948// CHECK9-NEXT: entry:949// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8950// CHECK9-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 8951// CHECK9-NEXT: [[T_INDIRECT_ADDR:%.*]] = alloca ptr, align 8952// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8953// CHECK9-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 8954// CHECK9-NEXT: store ptr [[T]], ptr [[T_INDIRECT_ADDR]], align 8955// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8956// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[S_ADDR]], align 8957// CHECK9-NEXT: call void @_ZN1SIiEC2ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[THIS1]], ptr nonnull align 4 dereferenceable(4) [[TMP0]], ptr [[T]])958// CHECK9-NEXT: ret void959//960//961// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiED1Ev962// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {963// CHECK9-NEXT: entry:964// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 8965// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 8966// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8967// CHECK9-NEXT: call void @_ZN1SIiED2Ev(ptr nonnull align 4 dereferenceable(4) [[THIS1]]) #[[ATTR4]]968// CHECK9-NEXT: ret void969//970//971// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l81972// CHECK9-SAME: (i64 [[T_VAR:%.*]]) #[[ATTR3]] {973// CHECK9-NEXT: entry:974// CHECK9-NEXT: [[T_VAR_ADDR:%.*]] = alloca i64, align 8975// CHECK9-NEXT: [[T_VAR_CASTED:%.*]] = alloca i64, align 8976// CHECK9-NEXT: store i64 [[T_VAR]], ptr [[T_VAR_ADDR]], align 8977// CHECK9-NEXT: [[TMP0:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 4978// CHECK9-NEXT: store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 4979// CHECK9-NEXT: [[TMP1:%.*]] = load i64, ptr [[T_VAR_CASTED]], align 8980// CHECK9-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l81.omp_outlined, i64 [[TMP1]])981// CHECK9-NEXT: ret void982//983//984// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l81.omp_outlined985// CHECK9-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i64 [[T_VAR:%.*]]) #[[ATTR3]] {986// CHECK9-NEXT: entry:987// CHECK9-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8988// CHECK9-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8989// CHECK9-NEXT: [[T_VAR_ADDR:%.*]] = alloca i64, align 8990// CHECK9-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8991// CHECK9-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8992// CHECK9-NEXT: store i64 [[T_VAR]], ptr [[T_VAR_ADDR]], align 8993// CHECK9-NEXT: ret void994//995//996// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC2Ev997// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {998// CHECK9-NEXT: entry:999// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 81000// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 81001// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 81002// CHECK9-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 01003// CHECK9-NEXT: [[TMP0:%.*]] = load volatile i32, ptr @g, align 1281004// CHECK9-NEXT: store i32 [[TMP0]], ptr [[F]], align 41005// CHECK9-NEXT: ret void1006//1007//1008// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC2Ei1009// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], i32 signext [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat {1010// CHECK9-NEXT: entry:1011// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 81012// CHECK9-NEXT: [[A_ADDR:%.*]] = alloca i32, align 41013// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 81014// CHECK9-NEXT: store i32 [[A]], ptr [[A_ADDR]], align 41015// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 81016// CHECK9-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 01017// CHECK9-NEXT: [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41018// CHECK9-NEXT: [[TMP1:%.*]] = load volatile i32, ptr @g, align 1281019// CHECK9-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP0]], [[TMP1]]1020// CHECK9-NEXT: store i32 [[ADD]], ptr [[F]], align 41021// CHECK9-NEXT: ret void1022//1023//1024// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC2ERKS0_2St1025// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], ptr nonnull align 4 dereferenceable(4) [[S:%.*]], ptr [[T:%.*]]) unnamed_addr #[[ATTR1]] comdat {1026// CHECK9-NEXT: entry:1027// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 81028// CHECK9-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 81029// CHECK9-NEXT: [[T_INDIRECT_ADDR:%.*]] = alloca ptr, align 81030// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 81031// CHECK9-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 81032// CHECK9-NEXT: store ptr [[T]], ptr [[T_INDIRECT_ADDR]], align 81033// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 81034// CHECK9-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 01035// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[S_ADDR]], align 8, !nonnull [[META11]], !align [[META12]]1036// CHECK9-NEXT: [[F2:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0]], ptr [[TMP0]], i32 0, i32 01037// CHECK9-NEXT: [[TMP1:%.*]] = load i32, ptr [[F2]], align 41038// CHECK9-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[T]], i32 0, i32 01039// CHECK9-NEXT: [[TMP2:%.*]] = load i32, ptr [[A]], align 41040// CHECK9-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP1]], [[TMP2]]1041// CHECK9-NEXT: store i32 [[ADD]], ptr [[F]], align 41042// CHECK9-NEXT: ret void1043//1044//1045// CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiED2Ev1046// CHECK9-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {1047// CHECK9-NEXT: entry:1048// CHECK9-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 81049// CHECK9-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 81050// CHECK9-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 81051// CHECK9-NEXT: ret void1052//1053//1054// CHECK11-LABEL: define {{[^@]+}}@main1055// CHECK11-SAME: () #[[ATTR0:[0-9]+]] {1056// CHECK11-NEXT: entry:1057// CHECK11-NEXT: [[RETVAL:%.*]] = alloca i32, align 41058// CHECK11-NEXT: [[TEST:%.*]] = alloca [[STRUCT_S:%.*]], align 41059// CHECK11-NEXT: [[T_VAR:%.*]] = alloca i32, align 41060// CHECK11-NEXT: [[VEC:%.*]] = alloca [2 x i32], align 41061// CHECK11-NEXT: [[S_ARR:%.*]] = alloca [2 x %struct.S], align 41062// CHECK11-NEXT: [[VAR:%.*]] = alloca [[STRUCT_S]], align 41063// CHECK11-NEXT: [[T_VAR_CASTED:%.*]] = alloca i32, align 41064// CHECK11-NEXT: [[SIVAR_CASTED:%.*]] = alloca i32, align 41065// CHECK11-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [5 x ptr], align 41066// CHECK11-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [5 x ptr], align 41067// CHECK11-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [5 x ptr], align 41068// CHECK11-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 81069// CHECK11-NEXT: [[T_VAR_CASTED1:%.*]] = alloca i32, align 41070// CHECK11-NEXT: [[DOTOFFLOAD_BASEPTRS2:%.*]] = alloca [1 x ptr], align 41071// CHECK11-NEXT: [[DOTOFFLOAD_PTRS3:%.*]] = alloca [1 x ptr], align 41072// CHECK11-NEXT: [[DOTOFFLOAD_MAPPERS4:%.*]] = alloca [1 x ptr], align 41073// CHECK11-NEXT: [[KERNEL_ARGS5:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 81074// CHECK11-NEXT: store i32 0, ptr [[RETVAL]], align 41075// CHECK11-NEXT: call void @_ZN1SIfEC1Ev(ptr nonnull align 4 dereferenceable(4) [[TEST]])1076// CHECK11-NEXT: store i32 0, ptr [[T_VAR]], align 41077// CHECK11-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[VEC]], ptr align 4 @__const.main.vec, i32 8, i1 false)1078// CHECK11-NEXT: call void @_ZN1SIfEC1Ef(ptr nonnull align 4 dereferenceable(4) [[S_ARR]], float 1.000000e+00)1079// CHECK11-NEXT: [[ARRAYINIT_ELEMENT:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[S_ARR]], i32 11080// CHECK11-NEXT: call void @_ZN1SIfEC1Ef(ptr nonnull align 4 dereferenceable(4) [[ARRAYINIT_ELEMENT]], float 2.000000e+00)1081// CHECK11-NEXT: call void @_ZN1SIfEC1Ef(ptr nonnull align 4 dereferenceable(4) [[VAR]], float 3.000000e+00)1082// CHECK11-NEXT: [[TMP0:%.*]] = load i32, ptr [[T_VAR]], align 41083// CHECK11-NEXT: store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 41084// CHECK11-NEXT: [[TMP1:%.*]] = load i32, ptr [[T_VAR_CASTED]], align 41085// CHECK11-NEXT: [[TMP2:%.*]] = load i32, ptr @_ZZ4mainE5sivar, align 41086// CHECK11-NEXT: store i32 [[TMP2]], ptr [[SIVAR_CASTED]], align 41087// CHECK11-NEXT: [[TMP3:%.*]] = load i32, ptr [[SIVAR_CASTED]], align 41088// CHECK11-NEXT: [[TMP4:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01089// CHECK11-NEXT: store i32 [[TMP1]], ptr [[TMP4]], align 41090// CHECK11-NEXT: [[TMP5:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01091// CHECK11-NEXT: store i32 [[TMP1]], ptr [[TMP5]], align 41092// CHECK11-NEXT: [[TMP6:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 01093// CHECK11-NEXT: store ptr null, ptr [[TMP6]], align 41094// CHECK11-NEXT: [[TMP7:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 11095// CHECK11-NEXT: store ptr [[VEC]], ptr [[TMP7]], align 41096// CHECK11-NEXT: [[TMP8:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 11097// CHECK11-NEXT: store ptr [[VEC]], ptr [[TMP8]], align 41098// CHECK11-NEXT: [[TMP9:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 11099// CHECK11-NEXT: store ptr null, ptr [[TMP9]], align 41100// CHECK11-NEXT: [[TMP10:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 21101// CHECK11-NEXT: store ptr [[S_ARR]], ptr [[TMP10]], align 41102// CHECK11-NEXT: [[TMP11:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 21103// CHECK11-NEXT: store ptr [[S_ARR]], ptr [[TMP11]], align 41104// CHECK11-NEXT: [[TMP12:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 21105// CHECK11-NEXT: store ptr null, ptr [[TMP12]], align 41106// CHECK11-NEXT: [[TMP13:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 31107// CHECK11-NEXT: store ptr [[VAR]], ptr [[TMP13]], align 41108// CHECK11-NEXT: [[TMP14:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 31109// CHECK11-NEXT: store ptr [[VAR]], ptr [[TMP14]], align 41110// CHECK11-NEXT: [[TMP15:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 31111// CHECK11-NEXT: store ptr null, ptr [[TMP15]], align 41112// CHECK11-NEXT: [[TMP16:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 41113// CHECK11-NEXT: store i32 [[TMP3]], ptr [[TMP16]], align 41114// CHECK11-NEXT: [[TMP17:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 41115// CHECK11-NEXT: store i32 [[TMP3]], ptr [[TMP17]], align 41116// CHECK11-NEXT: [[TMP18:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 41117// CHECK11-NEXT: store ptr null, ptr [[TMP18]], align 41118// CHECK11-NEXT: [[TMP19:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01119// CHECK11-NEXT: [[TMP20:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01120// CHECK11-NEXT: [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 01121// CHECK11-NEXT: store i32 3, ptr [[TMP21]], align 41122// CHECK11-NEXT: [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11123// CHECK11-NEXT: store i32 5, ptr [[TMP22]], align 41124// CHECK11-NEXT: [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21125// CHECK11-NEXT: store ptr [[TMP19]], ptr [[TMP23]], align 41126// CHECK11-NEXT: [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31127// CHECK11-NEXT: store ptr [[TMP20]], ptr [[TMP24]], align 41128// CHECK11-NEXT: [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41129// CHECK11-NEXT: store ptr @.offload_sizes, ptr [[TMP25]], align 41130// CHECK11-NEXT: [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51131// CHECK11-NEXT: store ptr @.offload_maptypes, ptr [[TMP26]], align 41132// CHECK11-NEXT: [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61133// CHECK11-NEXT: store ptr null, ptr [[TMP27]], align 41134// CHECK11-NEXT: [[TMP28:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71135// CHECK11-NEXT: store ptr null, ptr [[TMP28]], align 41136// CHECK11-NEXT: [[TMP29:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81137// CHECK11-NEXT: store i64 0, ptr [[TMP29]], align 81138// CHECK11-NEXT: [[TMP30:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91139// CHECK11-NEXT: store i64 0, ptr [[TMP30]], align 81140// CHECK11-NEXT: [[TMP31:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101141// CHECK11-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP31]], align 41142// CHECK11-NEXT: [[TMP32:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111143// CHECK11-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP32]], align 41144// CHECK11-NEXT: [[TMP33:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121145// CHECK11-NEXT: store i32 0, ptr [[TMP33]], align 41146// CHECK11-NEXT: [[TMP34:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l109.region_id, ptr [[KERNEL_ARGS]])1147// CHECK11-NEXT: [[TMP35:%.*]] = icmp ne i32 [[TMP34]], 01148// CHECK11-NEXT: br i1 [[TMP35]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]1149// CHECK11: omp_offload.failed:1150// CHECK11-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l109(i32 [[TMP1]], ptr [[VEC]], ptr [[S_ARR]], ptr [[VAR]], i32 [[TMP3]]) #[[ATTR4:[0-9]+]]1151// CHECK11-NEXT: br label [[OMP_OFFLOAD_CONT]]1152// CHECK11: omp_offload.cont:1153// CHECK11-NEXT: [[TMP36:%.*]] = load i32, ptr [[T_VAR]], align 41154// CHECK11-NEXT: store i32 [[TMP36]], ptr [[T_VAR_CASTED1]], align 41155// CHECK11-NEXT: [[TMP37:%.*]] = load i32, ptr [[T_VAR_CASTED1]], align 41156// CHECK11-NEXT: [[TMP38:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS2]], i32 0, i32 01157// CHECK11-NEXT: store i32 [[TMP37]], ptr [[TMP38]], align 41158// CHECK11-NEXT: [[TMP39:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS3]], i32 0, i32 01159// CHECK11-NEXT: store i32 [[TMP37]], ptr [[TMP39]], align 41160// CHECK11-NEXT: [[TMP40:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS4]], i32 0, i32 01161// CHECK11-NEXT: store ptr null, ptr [[TMP40]], align 41162// CHECK11-NEXT: [[TMP41:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS2]], i32 0, i32 01163// CHECK11-NEXT: [[TMP42:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS3]], i32 0, i32 01164// CHECK11-NEXT: [[TMP43:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 01165// CHECK11-NEXT: store i32 3, ptr [[TMP43]], align 41166// CHECK11-NEXT: [[TMP44:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 11167// CHECK11-NEXT: store i32 1, ptr [[TMP44]], align 41168// CHECK11-NEXT: [[TMP45:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 21169// CHECK11-NEXT: store ptr [[TMP41]], ptr [[TMP45]], align 41170// CHECK11-NEXT: [[TMP46:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 31171// CHECK11-NEXT: store ptr [[TMP42]], ptr [[TMP46]], align 41172// CHECK11-NEXT: [[TMP47:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 41173// CHECK11-NEXT: store ptr @.offload_sizes.1, ptr [[TMP47]], align 41174// CHECK11-NEXT: [[TMP48:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 51175// CHECK11-NEXT: store ptr @.offload_maptypes.2, ptr [[TMP48]], align 41176// CHECK11-NEXT: [[TMP49:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 61177// CHECK11-NEXT: store ptr null, ptr [[TMP49]], align 41178// CHECK11-NEXT: [[TMP50:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 71179// CHECK11-NEXT: store ptr null, ptr [[TMP50]], align 41180// CHECK11-NEXT: [[TMP51:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 81181// CHECK11-NEXT: store i64 0, ptr [[TMP51]], align 81182// CHECK11-NEXT: [[TMP52:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 91183// CHECK11-NEXT: store i64 0, ptr [[TMP52]], align 81184// CHECK11-NEXT: [[TMP53:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 101185// CHECK11-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP53]], align 41186// CHECK11-NEXT: [[TMP54:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 111187// CHECK11-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP54]], align 41188// CHECK11-NEXT: [[TMP55:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 121189// CHECK11-NEXT: store i32 0, ptr [[TMP55]], align 41190// CHECK11-NEXT: [[TMP56:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l116.region_id, ptr [[KERNEL_ARGS5]])1191// CHECK11-NEXT: [[TMP57:%.*]] = icmp ne i32 [[TMP56]], 01192// CHECK11-NEXT: br i1 [[TMP57]], label [[OMP_OFFLOAD_FAILED6:%.*]], label [[OMP_OFFLOAD_CONT7:%.*]]1193// CHECK11: omp_offload.failed6:1194// CHECK11-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l116(i32 [[TMP37]]) #[[ATTR4]]1195// CHECK11-NEXT: br label [[OMP_OFFLOAD_CONT7]]1196// CHECK11: omp_offload.cont7:1197// CHECK11-NEXT: [[CALL:%.*]] = call i32 @_Z5tmainIiET_v()1198// CHECK11-NEXT: store i32 [[CALL]], ptr [[RETVAL]], align 41199// CHECK11-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[VAR]]) #[[ATTR4]]1200// CHECK11-NEXT: [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i32 0, i32 01201// CHECK11-NEXT: [[TMP58:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN]], i32 21202// CHECK11-NEXT: br label [[ARRAYDESTROY_BODY:%.*]]1203// CHECK11: arraydestroy.body:1204// CHECK11-NEXT: [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP58]], [[OMP_OFFLOAD_CONT7]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]1205// CHECK11-NEXT: [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i32 -11206// CHECK11-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]1207// CHECK11-NEXT: [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN]]1208// CHECK11-NEXT: br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]1209// CHECK11: arraydestroy.done8:1210// CHECK11-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[TEST]]) #[[ATTR4]]1211// CHECK11-NEXT: [[TMP59:%.*]] = load i32, ptr [[RETVAL]], align 41212// CHECK11-NEXT: ret i32 [[TMP59]]1213//1214//1215// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC1Ev1216// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1:[0-9]+]] comdat align 2 {1217// CHECK11-NEXT: entry:1218// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41219// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41220// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41221// CHECK11-NEXT: call void @_ZN1SIfEC2Ev(ptr nonnull align 4 dereferenceable(4) [[THIS1]])1222// CHECK11-NEXT: ret void1223//1224//1225// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC1Ef1226// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], float [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1227// CHECK11-NEXT: entry:1228// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41229// CHECK11-NEXT: [[A_ADDR:%.*]] = alloca float, align 41230// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41231// CHECK11-NEXT: store float [[A]], ptr [[A_ADDR]], align 41232// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41233// CHECK11-NEXT: [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 41234// CHECK11-NEXT: call void @_ZN1SIfEC2Ef(ptr nonnull align 4 dereferenceable(4) [[THIS1]], float [[TMP0]])1235// CHECK11-NEXT: ret void1236//1237//1238// CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l1091239// CHECK11-SAME: (i32 [[T_VAR:%.*]], ptr nonnull align 4 dereferenceable(8) [[VEC:%.*]], ptr nonnull align 4 dereferenceable(8) [[S_ARR:%.*]], ptr nonnull align 4 dereferenceable(4) [[VAR:%.*]], i32 [[SIVAR:%.*]]) #[[ATTR3:[0-9]+]] {1240// CHECK11-NEXT: entry:1241// CHECK11-NEXT: [[T_VAR_ADDR:%.*]] = alloca i32, align 41242// CHECK11-NEXT: [[VEC_ADDR:%.*]] = alloca ptr, align 41243// CHECK11-NEXT: [[S_ARR_ADDR:%.*]] = alloca ptr, align 41244// CHECK11-NEXT: [[VAR_ADDR:%.*]] = alloca ptr, align 41245// CHECK11-NEXT: [[SIVAR_ADDR:%.*]] = alloca i32, align 41246// CHECK11-NEXT: [[T_VAR_CASTED:%.*]] = alloca i32, align 41247// CHECK11-NEXT: [[SIVAR_CASTED:%.*]] = alloca i32, align 41248// CHECK11-NEXT: store i32 [[T_VAR]], ptr [[T_VAR_ADDR]], align 41249// CHECK11-NEXT: store ptr [[VEC]], ptr [[VEC_ADDR]], align 41250// CHECK11-NEXT: store ptr [[S_ARR]], ptr [[S_ARR_ADDR]], align 41251// CHECK11-NEXT: store ptr [[VAR]], ptr [[VAR_ADDR]], align 41252// CHECK11-NEXT: store i32 [[SIVAR]], ptr [[SIVAR_ADDR]], align 41253// CHECK11-NEXT: [[TMP0:%.*]] = load ptr, ptr [[VEC_ADDR]], align 4, !nonnull [[META12:![0-9]+]], !align [[META13:![0-9]+]]1254// CHECK11-NEXT: [[TMP1:%.*]] = load ptr, ptr [[S_ARR_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1255// CHECK11-NEXT: [[TMP2:%.*]] = load ptr, ptr [[VAR_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1256// CHECK11-NEXT: [[TMP3:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 41257// CHECK11-NEXT: store i32 [[TMP3]], ptr [[T_VAR_CASTED]], align 41258// CHECK11-NEXT: [[TMP4:%.*]] = load i32, ptr [[T_VAR_CASTED]], align 41259// CHECK11-NEXT: [[TMP5:%.*]] = load i32, ptr [[SIVAR_ADDR]], align 41260// CHECK11-NEXT: store i32 [[TMP5]], ptr [[SIVAR_CASTED]], align 41261// CHECK11-NEXT: [[TMP6:%.*]] = load i32, ptr [[SIVAR_CASTED]], align 41262// CHECK11-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 5, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l109.omp_outlined, ptr [[TMP0]], i32 [[TMP4]], ptr [[TMP1]], ptr [[TMP2]], i32 [[TMP6]])1263// CHECK11-NEXT: ret void1264//1265//1266// CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l109.omp_outlined1267// CHECK11-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], ptr nonnull align 4 dereferenceable(8) [[VEC:%.*]], i32 [[T_VAR:%.*]], ptr nonnull align 4 dereferenceable(8) [[S_ARR:%.*]], ptr nonnull align 4 dereferenceable(4) [[VAR:%.*]], i32 [[SIVAR:%.*]]) #[[ATTR3]] {1268// CHECK11-NEXT: entry:1269// CHECK11-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 41270// CHECK11-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 41271// CHECK11-NEXT: [[VEC_ADDR:%.*]] = alloca ptr, align 41272// CHECK11-NEXT: [[T_VAR_ADDR:%.*]] = alloca i32, align 41273// CHECK11-NEXT: [[S_ARR_ADDR:%.*]] = alloca ptr, align 41274// CHECK11-NEXT: [[VAR_ADDR:%.*]] = alloca ptr, align 41275// CHECK11-NEXT: [[SIVAR_ADDR:%.*]] = alloca i32, align 41276// CHECK11-NEXT: [[VEC1:%.*]] = alloca [2 x i32], align 41277// CHECK11-NEXT: [[S_ARR2:%.*]] = alloca [2 x %struct.S], align 41278// CHECK11-NEXT: [[AGG_TMP:%.*]] = alloca [[STRUCT_ST:%.*]], align 41279// CHECK11-NEXT: [[VAR4:%.*]] = alloca [[STRUCT_S:%.*]], align 41280// CHECK11-NEXT: [[AGG_TMP5:%.*]] = alloca [[STRUCT_ST]], align 41281// CHECK11-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 41282// CHECK11-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 41283// CHECK11-NEXT: store ptr [[VEC]], ptr [[VEC_ADDR]], align 41284// CHECK11-NEXT: store i32 [[T_VAR]], ptr [[T_VAR_ADDR]], align 41285// CHECK11-NEXT: store ptr [[S_ARR]], ptr [[S_ARR_ADDR]], align 41286// CHECK11-NEXT: store ptr [[VAR]], ptr [[VAR_ADDR]], align 41287// CHECK11-NEXT: store i32 [[SIVAR]], ptr [[SIVAR_ADDR]], align 41288// CHECK11-NEXT: [[TMP0:%.*]] = load ptr, ptr [[VEC_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1289// CHECK11-NEXT: [[TMP1:%.*]] = load ptr, ptr [[S_ARR_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1290// CHECK11-NEXT: [[TMP2:%.*]] = load ptr, ptr [[VAR_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1291// CHECK11-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[VEC1]], ptr align 4 [[TMP0]], i32 8, i1 false)1292// CHECK11-NEXT: [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR2]], i32 0, i32 01293// CHECK11-NEXT: [[TMP3:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN]], i32 21294// CHECK11-NEXT: [[OMP_ARRAYCPY_ISEMPTY:%.*]] = icmp eq ptr [[ARRAY_BEGIN]], [[TMP3]]1295// CHECK11-NEXT: br i1 [[OMP_ARRAYCPY_ISEMPTY]], label [[OMP_ARRAYCPY_DONE3:%.*]], label [[OMP_ARRAYCPY_BODY:%.*]]1296// CHECK11: omp.arraycpy.body:1297// CHECK11-NEXT: [[OMP_ARRAYCPY_SRCELEMENTPAST:%.*]] = phi ptr [ [[TMP1]], [[ENTRY:%.*]] ], [ [[OMP_ARRAYCPY_SRC_ELEMENT:%.*]], [[OMP_ARRAYCPY_BODY]] ]1298// CHECK11-NEXT: [[OMP_ARRAYCPY_DESTELEMENTPAST:%.*]] = phi ptr [ [[ARRAY_BEGIN]], [[ENTRY]] ], [ [[OMP_ARRAYCPY_DEST_ELEMENT:%.*]], [[OMP_ARRAYCPY_BODY]] ]1299// CHECK11-NEXT: call void @_ZN2StC1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP]])1300// CHECK11-NEXT: call void @_ZN1SIfEC1ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[OMP_ARRAYCPY_DESTELEMENTPAST]], ptr nonnull align 4 dereferenceable(4) [[OMP_ARRAYCPY_SRCELEMENTPAST]], ptr [[AGG_TMP]])1301// CHECK11-NEXT: call void @_ZN2StD1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP]]) #[[ATTR4]]1302// CHECK11-NEXT: [[OMP_ARRAYCPY_DEST_ELEMENT]] = getelementptr [[STRUCT_S]], ptr [[OMP_ARRAYCPY_DESTELEMENTPAST]], i32 11303// CHECK11-NEXT: [[OMP_ARRAYCPY_SRC_ELEMENT]] = getelementptr [[STRUCT_S]], ptr [[OMP_ARRAYCPY_SRCELEMENTPAST]], i32 11304// CHECK11-NEXT: [[OMP_ARRAYCPY_DONE:%.*]] = icmp eq ptr [[OMP_ARRAYCPY_DEST_ELEMENT]], [[TMP3]]1305// CHECK11-NEXT: br i1 [[OMP_ARRAYCPY_DONE]], label [[OMP_ARRAYCPY_DONE3]], label [[OMP_ARRAYCPY_BODY]]1306// CHECK11: omp.arraycpy.done3:1307// CHECK11-NEXT: call void @_ZN2StC1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP5]])1308// CHECK11-NEXT: call void @_ZN1SIfEC1ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[VAR4]], ptr nonnull align 4 dereferenceable(4) [[TMP2]], ptr [[AGG_TMP5]])1309// CHECK11-NEXT: call void @_ZN2StD1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP5]]) #[[ATTR4]]1310// CHECK11-NEXT: [[TMP4:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 41311// CHECK11-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[VEC1]], i32 0, i32 01312// CHECK11-NEXT: store i32 [[TMP4]], ptr [[ARRAYIDX]], align 41313// CHECK11-NEXT: [[ARRAYIDX6:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR2]], i32 0, i32 01314// CHECK11-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[ARRAYIDX6]], ptr align 4 [[VAR4]], i32 4, i1 false)1315// CHECK11-NEXT: store i32 2, ptr [[SIVAR_ADDR]], align 41316// CHECK11-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[VAR4]]) #[[ATTR4]]1317// CHECK11-NEXT: [[ARRAY_BEGIN7:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR2]], i32 0, i32 01318// CHECK11-NEXT: [[TMP5:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN7]], i32 21319// CHECK11-NEXT: br label [[ARRAYDESTROY_BODY:%.*]]1320// CHECK11: arraydestroy.body:1321// CHECK11-NEXT: [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP5]], [[OMP_ARRAYCPY_DONE3]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]1322// CHECK11-NEXT: [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i32 -11323// CHECK11-NEXT: call void @_ZN1SIfED1Ev(ptr nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]1324// CHECK11-NEXT: [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN7]]1325// CHECK11-NEXT: br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]1326// CHECK11: arraydestroy.done8:1327// CHECK11-NEXT: ret void1328//1329//1330// CHECK11-LABEL: define {{[^@]+}}@_ZN2StC1Ev1331// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1332// CHECK11-NEXT: entry:1333// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41334// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41335// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41336// CHECK11-NEXT: call void @_ZN2StC2Ev(ptr nonnull align 4 dereferenceable(8) [[THIS1]])1337// CHECK11-NEXT: ret void1338//1339//1340// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC1ERKS0_2St1341// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], ptr nonnull align 4 dereferenceable(4) [[S:%.*]], ptr [[T:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1342// CHECK11-NEXT: entry:1343// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41344// CHECK11-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 41345// CHECK11-NEXT: [[T_INDIRECT_ADDR:%.*]] = alloca ptr, align 41346// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41347// CHECK11-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 41348// CHECK11-NEXT: store ptr [[T]], ptr [[T_INDIRECT_ADDR]], align 41349// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41350// CHECK11-NEXT: [[TMP0:%.*]] = load ptr, ptr [[S_ADDR]], align 41351// CHECK11-NEXT: call void @_ZN1SIfEC2ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[THIS1]], ptr nonnull align 4 dereferenceable(4) [[TMP0]], ptr [[T]])1352// CHECK11-NEXT: ret void1353//1354//1355// CHECK11-LABEL: define {{[^@]+}}@_ZN2StD1Ev1356// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1357// CHECK11-NEXT: entry:1358// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41359// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41360// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41361// CHECK11-NEXT: call void @_ZN2StD2Ev(ptr nonnull align 4 dereferenceable(8) [[THIS1]]) #[[ATTR4]]1362// CHECK11-NEXT: ret void1363//1364//1365// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfED1Ev1366// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1367// CHECK11-NEXT: entry:1368// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41369// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41370// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41371// CHECK11-NEXT: call void @_ZN1SIfED2Ev(ptr nonnull align 4 dereferenceable(4) [[THIS1]]) #[[ATTR4]]1372// CHECK11-NEXT: ret void1373//1374//1375// CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l1161376// CHECK11-SAME: (i32 [[T_VAR:%.*]]) #[[ATTR3]] {1377// CHECK11-NEXT: entry:1378// CHECK11-NEXT: [[T_VAR_ADDR:%.*]] = alloca i32, align 41379// CHECK11-NEXT: [[T_VAR_CASTED:%.*]] = alloca i32, align 41380// CHECK11-NEXT: store i32 [[T_VAR]], ptr [[T_VAR_ADDR]], align 41381// CHECK11-NEXT: [[TMP0:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 41382// CHECK11-NEXT: store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 41383// CHECK11-NEXT: [[TMP1:%.*]] = load i32, ptr [[T_VAR_CASTED]], align 41384// CHECK11-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l116.omp_outlined, i32 [[TMP1]])1385// CHECK11-NEXT: ret void1386//1387//1388// CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l116.omp_outlined1389// CHECK11-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i32 [[T_VAR:%.*]]) #[[ATTR3]] {1390// CHECK11-NEXT: entry:1391// CHECK11-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 41392// CHECK11-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 41393// CHECK11-NEXT: [[T_VAR_ADDR:%.*]] = alloca i32, align 41394// CHECK11-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 41395// CHECK11-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 41396// CHECK11-NEXT: store i32 [[T_VAR]], ptr [[T_VAR_ADDR]], align 41397// CHECK11-NEXT: ret void1398//1399//1400// CHECK11-LABEL: define {{[^@]+}}@_Z5tmainIiET_v1401// CHECK11-SAME: () #[[ATTR1]] comdat {1402// CHECK11-NEXT: entry:1403// CHECK11-NEXT: [[RETVAL:%.*]] = alloca i32, align 41404// CHECK11-NEXT: [[TEST:%.*]] = alloca [[STRUCT_S_0:%.*]], align 41405// CHECK11-NEXT: [[T_VAR:%.*]] = alloca i32, align 1281406// CHECK11-NEXT: [[VEC:%.*]] = alloca [2 x i32], align 1281407// CHECK11-NEXT: [[S_ARR:%.*]] = alloca [2 x %struct.S.0], align 1281408// CHECK11-NEXT: [[VAR:%.*]] = alloca [[STRUCT_S_0]], align 1281409// CHECK11-NEXT: [[T_VAR_CASTED:%.*]] = alloca i32, align 41410// CHECK11-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [4 x ptr], align 41411// CHECK11-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [4 x ptr], align 41412// CHECK11-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [4 x ptr], align 41413// CHECK11-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 81414// CHECK11-NEXT: [[T_VAR_CASTED1:%.*]] = alloca i32, align 41415// CHECK11-NEXT: [[DOTOFFLOAD_BASEPTRS2:%.*]] = alloca [1 x ptr], align 41416// CHECK11-NEXT: [[DOTOFFLOAD_PTRS3:%.*]] = alloca [1 x ptr], align 41417// CHECK11-NEXT: [[DOTOFFLOAD_MAPPERS4:%.*]] = alloca [1 x ptr], align 41418// CHECK11-NEXT: [[KERNEL_ARGS5:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 81419// CHECK11-NEXT: call void @_ZN1SIiEC1Ev(ptr nonnull align 4 dereferenceable(4) [[TEST]])1420// CHECK11-NEXT: store i32 0, ptr [[T_VAR]], align 1281421// CHECK11-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 128 [[VEC]], ptr align 128 @__const._Z5tmainIiET_v.vec, i32 8, i1 false)1422// CHECK11-NEXT: call void @_ZN1SIiEC1Ei(ptr nonnull align 4 dereferenceable(4) [[S_ARR]], i32 1)1423// CHECK11-NEXT: [[ARRAYINIT_ELEMENT:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[S_ARR]], i32 11424// CHECK11-NEXT: call void @_ZN1SIiEC1Ei(ptr nonnull align 4 dereferenceable(4) [[ARRAYINIT_ELEMENT]], i32 2)1425// CHECK11-NEXT: call void @_ZN1SIiEC1Ei(ptr nonnull align 4 dereferenceable(4) [[VAR]], i32 3)1426// CHECK11-NEXT: [[TMP0:%.*]] = load i32, ptr [[T_VAR]], align 1281427// CHECK11-NEXT: store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 41428// CHECK11-NEXT: [[TMP1:%.*]] = load i32, ptr [[T_VAR_CASTED]], align 41429// CHECK11-NEXT: [[TMP2:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01430// CHECK11-NEXT: store i32 [[TMP1]], ptr [[TMP2]], align 41431// CHECK11-NEXT: [[TMP3:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01432// CHECK11-NEXT: store i32 [[TMP1]], ptr [[TMP3]], align 41433// CHECK11-NEXT: [[TMP4:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 01434// CHECK11-NEXT: store ptr null, ptr [[TMP4]], align 41435// CHECK11-NEXT: [[TMP5:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 11436// CHECK11-NEXT: store ptr [[VEC]], ptr [[TMP5]], align 41437// CHECK11-NEXT: [[TMP6:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 11438// CHECK11-NEXT: store ptr [[VEC]], ptr [[TMP6]], align 41439// CHECK11-NEXT: [[TMP7:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 11440// CHECK11-NEXT: store ptr null, ptr [[TMP7]], align 41441// CHECK11-NEXT: [[TMP8:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 21442// CHECK11-NEXT: store ptr [[S_ARR]], ptr [[TMP8]], align 41443// CHECK11-NEXT: [[TMP9:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 21444// CHECK11-NEXT: store ptr [[S_ARR]], ptr [[TMP9]], align 41445// CHECK11-NEXT: [[TMP10:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 21446// CHECK11-NEXT: store ptr null, ptr [[TMP10]], align 41447// CHECK11-NEXT: [[TMP11:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 31448// CHECK11-NEXT: store ptr [[VAR]], ptr [[TMP11]], align 41449// CHECK11-NEXT: [[TMP12:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 31450// CHECK11-NEXT: store ptr [[VAR]], ptr [[TMP12]], align 41451// CHECK11-NEXT: [[TMP13:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 31452// CHECK11-NEXT: store ptr null, ptr [[TMP13]], align 41453// CHECK11-NEXT: [[TMP14:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01454// CHECK11-NEXT: [[TMP15:%.*]] = getelementptr inbounds [4 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01455// CHECK11-NEXT: [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 01456// CHECK11-NEXT: store i32 3, ptr [[TMP16]], align 41457// CHECK11-NEXT: [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11458// CHECK11-NEXT: store i32 4, ptr [[TMP17]], align 41459// CHECK11-NEXT: [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21460// CHECK11-NEXT: store ptr [[TMP14]], ptr [[TMP18]], align 41461// CHECK11-NEXT: [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31462// CHECK11-NEXT: store ptr [[TMP15]], ptr [[TMP19]], align 41463// CHECK11-NEXT: [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41464// CHECK11-NEXT: store ptr @.offload_sizes.3, ptr [[TMP20]], align 41465// CHECK11-NEXT: [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51466// CHECK11-NEXT: store ptr @.offload_maptypes.4, ptr [[TMP21]], align 41467// CHECK11-NEXT: [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61468// CHECK11-NEXT: store ptr null, ptr [[TMP22]], align 41469// CHECK11-NEXT: [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71470// CHECK11-NEXT: store ptr null, ptr [[TMP23]], align 41471// CHECK11-NEXT: [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81472// CHECK11-NEXT: store i64 0, ptr [[TMP24]], align 81473// CHECK11-NEXT: [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91474// CHECK11-NEXT: store i64 0, ptr [[TMP25]], align 81475// CHECK11-NEXT: [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101476// CHECK11-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP26]], align 41477// CHECK11-NEXT: [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111478// CHECK11-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP27]], align 41479// CHECK11-NEXT: [[TMP28:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121480// CHECK11-NEXT: store i32 0, ptr [[TMP28]], align 41481// CHECK11-NEXT: [[TMP29:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l75.region_id, ptr [[KERNEL_ARGS]])1482// CHECK11-NEXT: [[TMP30:%.*]] = icmp ne i32 [[TMP29]], 01483// CHECK11-NEXT: br i1 [[TMP30]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]1484// CHECK11: omp_offload.failed:1485// CHECK11-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l75(i32 [[TMP1]], ptr [[VEC]], ptr [[S_ARR]], ptr [[VAR]]) #[[ATTR4]]1486// CHECK11-NEXT: br label [[OMP_OFFLOAD_CONT]]1487// CHECK11: omp_offload.cont:1488// CHECK11-NEXT: [[TMP31:%.*]] = load i32, ptr [[T_VAR]], align 1281489// CHECK11-NEXT: store i32 [[TMP31]], ptr [[T_VAR_CASTED1]], align 41490// CHECK11-NEXT: [[TMP32:%.*]] = load i32, ptr [[T_VAR_CASTED1]], align 41491// CHECK11-NEXT: [[TMP33:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS2]], i32 0, i32 01492// CHECK11-NEXT: store i32 [[TMP32]], ptr [[TMP33]], align 41493// CHECK11-NEXT: [[TMP34:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS3]], i32 0, i32 01494// CHECK11-NEXT: store i32 [[TMP32]], ptr [[TMP34]], align 41495// CHECK11-NEXT: [[TMP35:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS4]], i32 0, i32 01496// CHECK11-NEXT: store ptr null, ptr [[TMP35]], align 41497// CHECK11-NEXT: [[TMP36:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS2]], i32 0, i32 01498// CHECK11-NEXT: [[TMP37:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS3]], i32 0, i32 01499// CHECK11-NEXT: [[TMP38:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 01500// CHECK11-NEXT: store i32 3, ptr [[TMP38]], align 41501// CHECK11-NEXT: [[TMP39:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 11502// CHECK11-NEXT: store i32 1, ptr [[TMP39]], align 41503// CHECK11-NEXT: [[TMP40:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 21504// CHECK11-NEXT: store ptr [[TMP36]], ptr [[TMP40]], align 41505// CHECK11-NEXT: [[TMP41:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 31506// CHECK11-NEXT: store ptr [[TMP37]], ptr [[TMP41]], align 41507// CHECK11-NEXT: [[TMP42:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 41508// CHECK11-NEXT: store ptr @.offload_sizes.5, ptr [[TMP42]], align 41509// CHECK11-NEXT: [[TMP43:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 51510// CHECK11-NEXT: store ptr @.offload_maptypes.6, ptr [[TMP43]], align 41511// CHECK11-NEXT: [[TMP44:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 61512// CHECK11-NEXT: store ptr null, ptr [[TMP44]], align 41513// CHECK11-NEXT: [[TMP45:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 71514// CHECK11-NEXT: store ptr null, ptr [[TMP45]], align 41515// CHECK11-NEXT: [[TMP46:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 81516// CHECK11-NEXT: store i64 0, ptr [[TMP46]], align 81517// CHECK11-NEXT: [[TMP47:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 91518// CHECK11-NEXT: store i64 0, ptr [[TMP47]], align 81519// CHECK11-NEXT: [[TMP48:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 101520// CHECK11-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP48]], align 41521// CHECK11-NEXT: [[TMP49:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 111522// CHECK11-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP49]], align 41523// CHECK11-NEXT: [[TMP50:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS5]], i32 0, i32 121524// CHECK11-NEXT: store i32 0, ptr [[TMP50]], align 41525// CHECK11-NEXT: [[TMP51:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l81.region_id, ptr [[KERNEL_ARGS5]])1526// CHECK11-NEXT: [[TMP52:%.*]] = icmp ne i32 [[TMP51]], 01527// CHECK11-NEXT: br i1 [[TMP52]], label [[OMP_OFFLOAD_FAILED6:%.*]], label [[OMP_OFFLOAD_CONT7:%.*]]1528// CHECK11: omp_offload.failed6:1529// CHECK11-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l81(i32 [[TMP32]]) #[[ATTR4]]1530// CHECK11-NEXT: br label [[OMP_OFFLOAD_CONT7]]1531// CHECK11: omp_offload.cont7:1532// CHECK11-NEXT: store i32 0, ptr [[RETVAL]], align 41533// CHECK11-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[VAR]]) #[[ATTR4]]1534// CHECK11-NEXT: [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i32 0, i32 01535// CHECK11-NEXT: [[TMP53:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN]], i32 21536// CHECK11-NEXT: br label [[ARRAYDESTROY_BODY:%.*]]1537// CHECK11: arraydestroy.body:1538// CHECK11-NEXT: [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP53]], [[OMP_OFFLOAD_CONT7]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]1539// CHECK11-NEXT: [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i32 -11540// CHECK11-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]1541// CHECK11-NEXT: [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN]]1542// CHECK11-NEXT: br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]1543// CHECK11: arraydestroy.done8:1544// CHECK11-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[TEST]]) #[[ATTR4]]1545// CHECK11-NEXT: [[TMP54:%.*]] = load i32, ptr [[RETVAL]], align 41546// CHECK11-NEXT: ret i32 [[TMP54]]1547//1548//1549// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC2Ev1550// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1551// CHECK11-NEXT: entry:1552// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41553// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41554// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41555// CHECK11-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 01556// CHECK11-NEXT: [[TMP0:%.*]] = load volatile i32, ptr @g, align 1281557// CHECK11-NEXT: [[CONV:%.*]] = sitofp i32 [[TMP0]] to float1558// CHECK11-NEXT: store float [[CONV]], ptr [[F]], align 41559// CHECK11-NEXT: ret void1560//1561//1562// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC2Ef1563// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], float [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1564// CHECK11-NEXT: entry:1565// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41566// CHECK11-NEXT: [[A_ADDR:%.*]] = alloca float, align 41567// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41568// CHECK11-NEXT: store float [[A]], ptr [[A_ADDR]], align 41569// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41570// CHECK11-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 01571// CHECK11-NEXT: [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 41572// CHECK11-NEXT: [[TMP1:%.*]] = load volatile i32, ptr @g, align 1281573// CHECK11-NEXT: [[CONV:%.*]] = sitofp i32 [[TMP1]] to float1574// CHECK11-NEXT: [[ADD:%.*]] = fadd float [[TMP0]], [[CONV]]1575// CHECK11-NEXT: store float [[ADD]], ptr [[F]], align 41576// CHECK11-NEXT: ret void1577//1578//1579// CHECK11-LABEL: define {{[^@]+}}@_ZN2StC2Ev1580// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1581// CHECK11-NEXT: entry:1582// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41583// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41584// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41585// CHECK11-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[THIS1]], i32 0, i32 01586// CHECK11-NEXT: store i32 0, ptr [[A]], align 41587// CHECK11-NEXT: [[B:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[THIS1]], i32 0, i32 11588// CHECK11-NEXT: store i32 0, ptr [[B]], align 41589// CHECK11-NEXT: ret void1590//1591//1592// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC2ERKS0_2St1593// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], ptr nonnull align 4 dereferenceable(4) [[S:%.*]], ptr [[T:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1594// CHECK11-NEXT: entry:1595// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41596// CHECK11-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 41597// CHECK11-NEXT: [[T_INDIRECT_ADDR:%.*]] = alloca ptr, align 41598// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41599// CHECK11-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 41600// CHECK11-NEXT: store ptr [[T]], ptr [[T_INDIRECT_ADDR]], align 41601// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41602// CHECK11-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 01603// CHECK11-NEXT: [[TMP0:%.*]] = load ptr, ptr [[S_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1604// CHECK11-NEXT: [[F2:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[TMP0]], i32 0, i32 01605// CHECK11-NEXT: [[TMP1:%.*]] = load float, ptr [[F2]], align 41606// CHECK11-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[T]], i32 0, i32 01607// CHECK11-NEXT: [[TMP2:%.*]] = load i32, ptr [[A]], align 41608// CHECK11-NEXT: [[CONV:%.*]] = sitofp i32 [[TMP2]] to float1609// CHECK11-NEXT: [[ADD:%.*]] = fadd float [[TMP1]], [[CONV]]1610// CHECK11-NEXT: store float [[ADD]], ptr [[F]], align 41611// CHECK11-NEXT: ret void1612//1613//1614// CHECK11-LABEL: define {{[^@]+}}@_ZN2StD2Ev1615// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1616// CHECK11-NEXT: entry:1617// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41618// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41619// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41620// CHECK11-NEXT: ret void1621//1622//1623// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfED2Ev1624// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1625// CHECK11-NEXT: entry:1626// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41627// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41628// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41629// CHECK11-NEXT: ret void1630//1631//1632// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC1Ev1633// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1634// CHECK11-NEXT: entry:1635// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41636// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41637// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41638// CHECK11-NEXT: call void @_ZN1SIiEC2Ev(ptr nonnull align 4 dereferenceable(4) [[THIS1]])1639// CHECK11-NEXT: ret void1640//1641//1642// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC1Ei1643// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], i32 [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1644// CHECK11-NEXT: entry:1645// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41646// CHECK11-NEXT: [[A_ADDR:%.*]] = alloca i32, align 41647// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41648// CHECK11-NEXT: store i32 [[A]], ptr [[A_ADDR]], align 41649// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41650// CHECK11-NEXT: [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41651// CHECK11-NEXT: call void @_ZN1SIiEC2Ei(ptr nonnull align 4 dereferenceable(4) [[THIS1]], i32 [[TMP0]])1652// CHECK11-NEXT: ret void1653//1654//1655// CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l751656// CHECK11-SAME: (i32 [[T_VAR:%.*]], ptr nonnull align 4 dereferenceable(8) [[VEC:%.*]], ptr nonnull align 4 dereferenceable(8) [[S_ARR:%.*]], ptr nonnull align 4 dereferenceable(4) [[VAR:%.*]]) #[[ATTR3]] {1657// CHECK11-NEXT: entry:1658// CHECK11-NEXT: [[T_VAR_ADDR:%.*]] = alloca i32, align 41659// CHECK11-NEXT: [[VEC_ADDR:%.*]] = alloca ptr, align 41660// CHECK11-NEXT: [[S_ARR_ADDR:%.*]] = alloca ptr, align 41661// CHECK11-NEXT: [[VAR_ADDR:%.*]] = alloca ptr, align 41662// CHECK11-NEXT: [[T_VAR_CASTED:%.*]] = alloca i32, align 41663// CHECK11-NEXT: store i32 [[T_VAR]], ptr [[T_VAR_ADDR]], align 41664// CHECK11-NEXT: store ptr [[VEC]], ptr [[VEC_ADDR]], align 41665// CHECK11-NEXT: store ptr [[S_ARR]], ptr [[S_ARR_ADDR]], align 41666// CHECK11-NEXT: store ptr [[VAR]], ptr [[VAR_ADDR]], align 41667// CHECK11-NEXT: [[TMP0:%.*]] = load ptr, ptr [[VEC_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1668// CHECK11-NEXT: [[TMP1:%.*]] = load ptr, ptr [[S_ARR_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1669// CHECK11-NEXT: [[TMP2:%.*]] = load ptr, ptr [[VAR_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1670// CHECK11-NEXT: [[TMP3:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 41671// CHECK11-NEXT: store i32 [[TMP3]], ptr [[T_VAR_CASTED]], align 41672// CHECK11-NEXT: [[TMP4:%.*]] = load i32, ptr [[T_VAR_CASTED]], align 41673// CHECK11-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 4, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l75.omp_outlined, ptr [[TMP0]], i32 [[TMP4]], ptr [[TMP1]], ptr [[TMP2]])1674// CHECK11-NEXT: ret void1675//1676//1677// CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l75.omp_outlined1678// CHECK11-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], ptr nonnull align 4 dereferenceable(8) [[VEC:%.*]], i32 [[T_VAR:%.*]], ptr nonnull align 4 dereferenceable(8) [[S_ARR:%.*]], ptr nonnull align 4 dereferenceable(4) [[VAR:%.*]]) #[[ATTR3]] {1679// CHECK11-NEXT: entry:1680// CHECK11-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 41681// CHECK11-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 41682// CHECK11-NEXT: [[VEC_ADDR:%.*]] = alloca ptr, align 41683// CHECK11-NEXT: [[T_VAR_ADDR:%.*]] = alloca i32, align 41684// CHECK11-NEXT: [[S_ARR_ADDR:%.*]] = alloca ptr, align 41685// CHECK11-NEXT: [[VAR_ADDR:%.*]] = alloca ptr, align 41686// CHECK11-NEXT: [[VEC1:%.*]] = alloca [2 x i32], align 1281687// CHECK11-NEXT: [[S_ARR2:%.*]] = alloca [2 x %struct.S.0], align 1281688// CHECK11-NEXT: [[AGG_TMP:%.*]] = alloca [[STRUCT_ST:%.*]], align 41689// CHECK11-NEXT: [[VAR4:%.*]] = alloca [[STRUCT_S_0:%.*]], align 1281690// CHECK11-NEXT: [[AGG_TMP5:%.*]] = alloca [[STRUCT_ST]], align 41691// CHECK11-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 41692// CHECK11-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 41693// CHECK11-NEXT: store ptr [[VEC]], ptr [[VEC_ADDR]], align 41694// CHECK11-NEXT: store i32 [[T_VAR]], ptr [[T_VAR_ADDR]], align 41695// CHECK11-NEXT: store ptr [[S_ARR]], ptr [[S_ARR_ADDR]], align 41696// CHECK11-NEXT: store ptr [[VAR]], ptr [[VAR_ADDR]], align 41697// CHECK11-NEXT: [[TMP0:%.*]] = load ptr, ptr [[VEC_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1698// CHECK11-NEXT: [[TMP1:%.*]] = load ptr, ptr [[S_ARR_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1699// CHECK11-NEXT: [[TMP2:%.*]] = load ptr, ptr [[VAR_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1700// CHECK11-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 128 [[VEC1]], ptr align 128 [[TMP0]], i32 8, i1 false)1701// CHECK11-NEXT: [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR2]], i32 0, i32 01702// CHECK11-NEXT: [[TMP3:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN]], i32 21703// CHECK11-NEXT: [[OMP_ARRAYCPY_ISEMPTY:%.*]] = icmp eq ptr [[ARRAY_BEGIN]], [[TMP3]]1704// CHECK11-NEXT: br i1 [[OMP_ARRAYCPY_ISEMPTY]], label [[OMP_ARRAYCPY_DONE3:%.*]], label [[OMP_ARRAYCPY_BODY:%.*]]1705// CHECK11: omp.arraycpy.body:1706// CHECK11-NEXT: [[OMP_ARRAYCPY_SRCELEMENTPAST:%.*]] = phi ptr [ [[TMP1]], [[ENTRY:%.*]] ], [ [[OMP_ARRAYCPY_SRC_ELEMENT:%.*]], [[OMP_ARRAYCPY_BODY]] ]1707// CHECK11-NEXT: [[OMP_ARRAYCPY_DESTELEMENTPAST:%.*]] = phi ptr [ [[ARRAY_BEGIN]], [[ENTRY]] ], [ [[OMP_ARRAYCPY_DEST_ELEMENT:%.*]], [[OMP_ARRAYCPY_BODY]] ]1708// CHECK11-NEXT: call void @_ZN2StC1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP]])1709// CHECK11-NEXT: call void @_ZN1SIiEC1ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[OMP_ARRAYCPY_DESTELEMENTPAST]], ptr nonnull align 4 dereferenceable(4) [[OMP_ARRAYCPY_SRCELEMENTPAST]], ptr [[AGG_TMP]])1710// CHECK11-NEXT: call void @_ZN2StD1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP]]) #[[ATTR4]]1711// CHECK11-NEXT: [[OMP_ARRAYCPY_DEST_ELEMENT]] = getelementptr [[STRUCT_S_0]], ptr [[OMP_ARRAYCPY_DESTELEMENTPAST]], i32 11712// CHECK11-NEXT: [[OMP_ARRAYCPY_SRC_ELEMENT]] = getelementptr [[STRUCT_S_0]], ptr [[OMP_ARRAYCPY_SRCELEMENTPAST]], i32 11713// CHECK11-NEXT: [[OMP_ARRAYCPY_DONE:%.*]] = icmp eq ptr [[OMP_ARRAYCPY_DEST_ELEMENT]], [[TMP3]]1714// CHECK11-NEXT: br i1 [[OMP_ARRAYCPY_DONE]], label [[OMP_ARRAYCPY_DONE3]], label [[OMP_ARRAYCPY_BODY]]1715// CHECK11: omp.arraycpy.done3:1716// CHECK11-NEXT: call void @_ZN2StC1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP5]])1717// CHECK11-NEXT: call void @_ZN1SIiEC1ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[VAR4]], ptr nonnull align 4 dereferenceable(4) [[TMP2]], ptr [[AGG_TMP5]])1718// CHECK11-NEXT: call void @_ZN2StD1Ev(ptr nonnull align 4 dereferenceable(8) [[AGG_TMP5]]) #[[ATTR4]]1719// CHECK11-NEXT: [[TMP4:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 41720// CHECK11-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[VEC1]], i32 0, i32 01721// CHECK11-NEXT: store i32 [[TMP4]], ptr [[ARRAYIDX]], align 1281722// CHECK11-NEXT: [[ARRAYIDX6:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR2]], i32 0, i32 01723// CHECK11-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 128 [[ARRAYIDX6]], ptr align 128 [[VAR4]], i32 4, i1 false)1724// CHECK11-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[VAR4]]) #[[ATTR4]]1725// CHECK11-NEXT: [[ARRAY_BEGIN7:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR2]], i32 0, i32 01726// CHECK11-NEXT: [[TMP5:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN7]], i32 21727// CHECK11-NEXT: br label [[ARRAYDESTROY_BODY:%.*]]1728// CHECK11: arraydestroy.body:1729// CHECK11-NEXT: [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP5]], [[OMP_ARRAYCPY_DONE3]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]1730// CHECK11-NEXT: [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i32 -11731// CHECK11-NEXT: call void @_ZN1SIiED1Ev(ptr nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]1732// CHECK11-NEXT: [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN7]]1733// CHECK11-NEXT: br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]1734// CHECK11: arraydestroy.done8:1735// CHECK11-NEXT: ret void1736//1737//1738// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC1ERKS0_2St1739// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], ptr nonnull align 4 dereferenceable(4) [[S:%.*]], ptr [[T:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1740// CHECK11-NEXT: entry:1741// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41742// CHECK11-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 41743// CHECK11-NEXT: [[T_INDIRECT_ADDR:%.*]] = alloca ptr, align 41744// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41745// CHECK11-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 41746// CHECK11-NEXT: store ptr [[T]], ptr [[T_INDIRECT_ADDR]], align 41747// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41748// CHECK11-NEXT: [[TMP0:%.*]] = load ptr, ptr [[S_ADDR]], align 41749// CHECK11-NEXT: call void @_ZN1SIiEC2ERKS0_2St(ptr nonnull align 4 dereferenceable(4) [[THIS1]], ptr nonnull align 4 dereferenceable(4) [[TMP0]], ptr [[T]])1750// CHECK11-NEXT: ret void1751//1752//1753// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiED1Ev1754// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1755// CHECK11-NEXT: entry:1756// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41757// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41758// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41759// CHECK11-NEXT: call void @_ZN1SIiED2Ev(ptr nonnull align 4 dereferenceable(4) [[THIS1]]) #[[ATTR4]]1760// CHECK11-NEXT: ret void1761//1762//1763// CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l811764// CHECK11-SAME: (i32 [[T_VAR:%.*]]) #[[ATTR3]] {1765// CHECK11-NEXT: entry:1766// CHECK11-NEXT: [[T_VAR_ADDR:%.*]] = alloca i32, align 41767// CHECK11-NEXT: [[T_VAR_CASTED:%.*]] = alloca i32, align 41768// CHECK11-NEXT: store i32 [[T_VAR]], ptr [[T_VAR_ADDR]], align 41769// CHECK11-NEXT: [[TMP0:%.*]] = load i32, ptr [[T_VAR_ADDR]], align 41770// CHECK11-NEXT: store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 41771// CHECK11-NEXT: [[TMP1:%.*]] = load i32, ptr [[T_VAR_CASTED]], align 41772// CHECK11-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l81.omp_outlined, i32 [[TMP1]])1773// CHECK11-NEXT: ret void1774//1775//1776// CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l81.omp_outlined1777// CHECK11-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i32 [[T_VAR:%.*]]) #[[ATTR3]] {1778// CHECK11-NEXT: entry:1779// CHECK11-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 41780// CHECK11-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 41781// CHECK11-NEXT: [[T_VAR_ADDR:%.*]] = alloca i32, align 41782// CHECK11-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 41783// CHECK11-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 41784// CHECK11-NEXT: store i32 [[T_VAR]], ptr [[T_VAR_ADDR]], align 41785// CHECK11-NEXT: ret void1786//1787//1788// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC2Ev1789// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1790// CHECK11-NEXT: entry:1791// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41792// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41793// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41794// CHECK11-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 01795// CHECK11-NEXT: [[TMP0:%.*]] = load volatile i32, ptr @g, align 1281796// CHECK11-NEXT: store i32 [[TMP0]], ptr [[F]], align 41797// CHECK11-NEXT: ret void1798//1799//1800// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC2Ei1801// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], i32 [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1802// CHECK11-NEXT: entry:1803// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41804// CHECK11-NEXT: [[A_ADDR:%.*]] = alloca i32, align 41805// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41806// CHECK11-NEXT: store i32 [[A]], ptr [[A_ADDR]], align 41807// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41808// CHECK11-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 01809// CHECK11-NEXT: [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 41810// CHECK11-NEXT: [[TMP1:%.*]] = load volatile i32, ptr @g, align 1281811// CHECK11-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP0]], [[TMP1]]1812// CHECK11-NEXT: store i32 [[ADD]], ptr [[F]], align 41813// CHECK11-NEXT: ret void1814//1815//1816// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC2ERKS0_2St1817// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]], ptr nonnull align 4 dereferenceable(4) [[S:%.*]], ptr [[T:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1818// CHECK11-NEXT: entry:1819// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41820// CHECK11-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 41821// CHECK11-NEXT: [[T_INDIRECT_ADDR:%.*]] = alloca ptr, align 41822// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41823// CHECK11-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 41824// CHECK11-NEXT: store ptr [[T]], ptr [[T_INDIRECT_ADDR]], align 41825// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41826// CHECK11-NEXT: [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 01827// CHECK11-NEXT: [[TMP0:%.*]] = load ptr, ptr [[S_ADDR]], align 4, !nonnull [[META12]], !align [[META13]]1828// CHECK11-NEXT: [[F2:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0]], ptr [[TMP0]], i32 0, i32 01829// CHECK11-NEXT: [[TMP1:%.*]] = load i32, ptr [[F2]], align 41830// CHECK11-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[T]], i32 0, i32 01831// CHECK11-NEXT: [[TMP2:%.*]] = load i32, ptr [[A]], align 41832// CHECK11-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP1]], [[TMP2]]1833// CHECK11-NEXT: store i32 [[ADD]], ptr [[F]], align 41834// CHECK11-NEXT: ret void1835//1836//1837// CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiED2Ev1838// CHECK11-SAME: (ptr nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {1839// CHECK11-NEXT: entry:1840// CHECK11-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 41841// CHECK11-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 41842// CHECK11-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 41843// CHECK11-NEXT: ret void1844//1845//1846// CHECK17-LABEL: define {{[^@]+}}@_Z10array_funcPfP2StiPg1847// CHECK17-SAME: (ptr [[A:%.*]], ptr [[S:%.*]], i32 signext [[N:%.*]], ptr [[VLA1:%.*]]) #[[ATTR0:[0-9]+]] {1848// CHECK17-NEXT: entry:1849// CHECK17-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 81850// CHECK17-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 81851// CHECK17-NEXT: [[N_ADDR:%.*]] = alloca i32, align 41852// CHECK17-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 81853// CHECK17-NEXT: [[SAVED_STACK:%.*]] = alloca ptr, align 81854// CHECK17-NEXT: [[__VLA_EXPR0:%.*]] = alloca i64, align 81855// CHECK17-NEXT: [[__VLA_EXPR1:%.*]] = alloca i64, align 81856// CHECK17-NEXT: [[N_CASTED:%.*]] = alloca i64, align 81857// CHECK17-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [8 x ptr], align 81858// CHECK17-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [8 x ptr], align 81859// CHECK17-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [8 x ptr], align 81860// CHECK17-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [8 x i64], align 81861// CHECK17-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 81862// CHECK17-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 81863// CHECK17-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 81864// CHECK17-NEXT: store i32 [[N]], ptr [[N_ADDR]], align 41865// CHECK17-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 81866// CHECK17-NEXT: [[TMP0:%.*]] = load i32, ptr [[N_ADDR]], align 41867// CHECK17-NEXT: [[TMP1:%.*]] = zext i32 [[TMP0]] to i641868// CHECK17-NEXT: [[TMP2:%.*]] = load i32, ptr [[N_ADDR]], align 41869// CHECK17-NEXT: [[TMP3:%.*]] = zext i32 [[TMP2]] to i641870// CHECK17-NEXT: [[TMP4:%.*]] = load i32, ptr [[N_ADDR]], align 41871// CHECK17-NEXT: [[TMP5:%.*]] = zext i32 [[TMP4]] to i641872// CHECK17-NEXT: [[TMP6:%.*]] = call ptr @llvm.stacksave.p0()1873// CHECK17-NEXT: store ptr [[TMP6]], ptr [[SAVED_STACK]], align 81874// CHECK17-NEXT: [[TMP7:%.*]] = mul nuw i64 [[TMP3]], [[TMP5]]1875// CHECK17-NEXT: [[VLA:%.*]] = alloca double, i64 [[TMP7]], align 1281876// CHECK17-NEXT: store i64 [[TMP3]], ptr [[__VLA_EXPR0]], align 81877// CHECK17-NEXT: store i64 [[TMP5]], ptr [[__VLA_EXPR1]], align 81878// CHECK17-NEXT: [[TMP8:%.*]] = load ptr, ptr [[A_ADDR]], align 81879// CHECK17-NEXT: [[TMP9:%.*]] = load ptr, ptr [[S_ADDR]], align 81880// CHECK17-NEXT: [[TMP10:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 81881// CHECK17-NEXT: [[TMP11:%.*]] = load i32, ptr [[N_ADDR]], align 41882// CHECK17-NEXT: store i32 [[TMP11]], ptr [[N_CASTED]], align 41883// CHECK17-NEXT: [[TMP12:%.*]] = load i64, ptr [[N_CASTED]], align 81884// CHECK17-NEXT: [[TMP13:%.*]] = mul nuw i64 [[TMP3]], [[TMP5]]1885// CHECK17-NEXT: [[TMP14:%.*]] = mul nuw i64 [[TMP13]], 81886// CHECK17-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 8 [[DOTOFFLOAD_SIZES]], ptr align 8 @.offload_sizes, i64 64, i1 false)1887// CHECK17-NEXT: [[TMP15:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01888// CHECK17-NEXT: store ptr [[TMP8]], ptr [[TMP15]], align 81889// CHECK17-NEXT: [[TMP16:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01890// CHECK17-NEXT: store ptr [[TMP8]], ptr [[TMP16]], align 81891// CHECK17-NEXT: [[TMP17:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 01892// CHECK17-NEXT: store ptr null, ptr [[TMP17]], align 81893// CHECK17-NEXT: [[TMP18:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 11894// CHECK17-NEXT: store ptr [[TMP9]], ptr [[TMP18]], align 81895// CHECK17-NEXT: [[TMP19:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 11896// CHECK17-NEXT: store ptr [[TMP9]], ptr [[TMP19]], align 81897// CHECK17-NEXT: [[TMP20:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 11898// CHECK17-NEXT: store ptr null, ptr [[TMP20]], align 81899// CHECK17-NEXT: [[TMP21:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 21900// CHECK17-NEXT: store i64 [[TMP1]], ptr [[TMP21]], align 81901// CHECK17-NEXT: [[TMP22:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 21902// CHECK17-NEXT: store i64 [[TMP1]], ptr [[TMP22]], align 81903// CHECK17-NEXT: [[TMP23:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 21904// CHECK17-NEXT: store ptr null, ptr [[TMP23]], align 81905// CHECK17-NEXT: [[TMP24:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 31906// CHECK17-NEXT: store ptr [[TMP10]], ptr [[TMP24]], align 81907// CHECK17-NEXT: [[TMP25:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 31908// CHECK17-NEXT: store ptr [[TMP10]], ptr [[TMP25]], align 81909// CHECK17-NEXT: [[TMP26:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 31910// CHECK17-NEXT: store ptr null, ptr [[TMP26]], align 81911// CHECK17-NEXT: [[TMP27:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 41912// CHECK17-NEXT: store i64 [[TMP3]], ptr [[TMP27]], align 81913// CHECK17-NEXT: [[TMP28:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 41914// CHECK17-NEXT: store i64 [[TMP3]], ptr [[TMP28]], align 81915// CHECK17-NEXT: [[TMP29:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 41916// CHECK17-NEXT: store ptr null, ptr [[TMP29]], align 81917// CHECK17-NEXT: [[TMP30:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 51918// CHECK17-NEXT: store i64 [[TMP5]], ptr [[TMP30]], align 81919// CHECK17-NEXT: [[TMP31:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 51920// CHECK17-NEXT: store i64 [[TMP5]], ptr [[TMP31]], align 81921// CHECK17-NEXT: [[TMP32:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 51922// CHECK17-NEXT: store ptr null, ptr [[TMP32]], align 81923// CHECK17-NEXT: [[TMP33:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 61924// CHECK17-NEXT: store ptr [[VLA]], ptr [[TMP33]], align 81925// CHECK17-NEXT: [[TMP34:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 61926// CHECK17-NEXT: store ptr [[VLA]], ptr [[TMP34]], align 81927// CHECK17-NEXT: [[TMP35:%.*]] = getelementptr inbounds [8 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 61928// CHECK17-NEXT: store i64 [[TMP14]], ptr [[TMP35]], align 81929// CHECK17-NEXT: [[TMP36:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 61930// CHECK17-NEXT: store ptr null, ptr [[TMP36]], align 81931// CHECK17-NEXT: [[TMP37:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 71932// CHECK17-NEXT: store i64 [[TMP12]], ptr [[TMP37]], align 81933// CHECK17-NEXT: [[TMP38:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 71934// CHECK17-NEXT: store i64 [[TMP12]], ptr [[TMP38]], align 81935// CHECK17-NEXT: [[TMP39:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 71936// CHECK17-NEXT: store ptr null, ptr [[TMP39]], align 81937// CHECK17-NEXT: [[TMP40:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 01938// CHECK17-NEXT: [[TMP41:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 01939// CHECK17-NEXT: [[TMP42:%.*]] = getelementptr inbounds [8 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 01940// CHECK17-NEXT: [[TMP43:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 01941// CHECK17-NEXT: store i32 3, ptr [[TMP43]], align 41942// CHECK17-NEXT: [[TMP44:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11943// CHECK17-NEXT: store i32 8, ptr [[TMP44]], align 41944// CHECK17-NEXT: [[TMP45:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 21945// CHECK17-NEXT: store ptr [[TMP40]], ptr [[TMP45]], align 81946// CHECK17-NEXT: [[TMP46:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31947// CHECK17-NEXT: store ptr [[TMP41]], ptr [[TMP46]], align 81948// CHECK17-NEXT: [[TMP47:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41949// CHECK17-NEXT: store ptr [[TMP42]], ptr [[TMP47]], align 81950// CHECK17-NEXT: [[TMP48:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51951// CHECK17-NEXT: store ptr @.offload_maptypes, ptr [[TMP48]], align 81952// CHECK17-NEXT: [[TMP49:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61953// CHECK17-NEXT: store ptr null, ptr [[TMP49]], align 81954// CHECK17-NEXT: [[TMP50:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71955// CHECK17-NEXT: store ptr null, ptr [[TMP50]], align 81956// CHECK17-NEXT: [[TMP51:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81957// CHECK17-NEXT: store i64 0, ptr [[TMP51]], align 81958// CHECK17-NEXT: [[TMP52:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91959// CHECK17-NEXT: store i64 0, ptr [[TMP52]], align 81960// CHECK17-NEXT: [[TMP53:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101961// CHECK17-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP53]], align 41962// CHECK17-NEXT: [[TMP54:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111963// CHECK17-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP54]], align 41964// CHECK17-NEXT: [[TMP55:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121965// CHECK17-NEXT: store i32 0, ptr [[TMP55]], align 41966// CHECK17-NEXT: [[TMP56:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPg_l152.region_id, ptr [[KERNEL_ARGS]])1967// CHECK17-NEXT: [[TMP57:%.*]] = icmp ne i32 [[TMP56]], 01968// CHECK17-NEXT: br i1 [[TMP57]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]1969// CHECK17: omp_offload.failed:1970// CHECK17-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPg_l152(ptr [[TMP8]], ptr [[TMP9]], i64 [[TMP1]], ptr [[TMP10]], i64 [[TMP3]], i64 [[TMP5]], ptr [[VLA]], i64 [[TMP12]]) #[[ATTR4:[0-9]+]]1971// CHECK17-NEXT: br label [[OMP_OFFLOAD_CONT]]1972// CHECK17: omp_offload.cont:1973// CHECK17-NEXT: [[TMP58:%.*]] = load ptr, ptr [[SAVED_STACK]], align 81974// CHECK17-NEXT: call void @llvm.stackrestore.p0(ptr [[TMP58]])1975// CHECK17-NEXT: ret void1976//1977//1978// CHECK17-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPg_l1521979// CHECK17-SAME: (ptr [[A:%.*]], ptr [[S:%.*]], i64 [[VLA:%.*]], ptr [[VLA1:%.*]], i64 [[VLA2:%.*]], i64 [[VLA4:%.*]], ptr nonnull align 8 dereferenceable(8) [[VLA26:%.*]], i64 [[N:%.*]]) #[[ATTR2:[0-9]+]] {1980// CHECK17-NEXT: entry:1981// CHECK17-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 81982// CHECK17-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 81983// CHECK17-NEXT: [[VLA_ADDR:%.*]] = alloca i64, align 81984// CHECK17-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 81985// CHECK17-NEXT: [[VLA_ADDR3:%.*]] = alloca i64, align 81986// CHECK17-NEXT: [[VLA_ADDR5:%.*]] = alloca i64, align 81987// CHECK17-NEXT: [[VLA2_ADDR:%.*]] = alloca ptr, align 81988// CHECK17-NEXT: [[N_ADDR:%.*]] = alloca i64, align 81989// CHECK17-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 81990// CHECK17-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 81991// CHECK17-NEXT: store i64 [[VLA]], ptr [[VLA_ADDR]], align 81992// CHECK17-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 81993// CHECK17-NEXT: store i64 [[VLA2]], ptr [[VLA_ADDR3]], align 81994// CHECK17-NEXT: store i64 [[VLA4]], ptr [[VLA_ADDR5]], align 81995// CHECK17-NEXT: store ptr [[VLA26]], ptr [[VLA2_ADDR]], align 81996// CHECK17-NEXT: store i64 [[N]], ptr [[N_ADDR]], align 81997// CHECK17-NEXT: [[TMP0:%.*]] = load i64, ptr [[VLA_ADDR]], align 81998// CHECK17-NEXT: [[TMP1:%.*]] = load i64, ptr [[VLA_ADDR3]], align 81999// CHECK17-NEXT: [[TMP2:%.*]] = load i64, ptr [[VLA_ADDR5]], align 82000// CHECK17-NEXT: [[TMP3:%.*]] = load ptr, ptr [[VLA2_ADDR]], align 8, !nonnull [[META8:![0-9]+]], !align [[META9:![0-9]+]]2001// CHECK17-NEXT: [[TMP4:%.*]] = load ptr, ptr [[S_ADDR]], align 82002// CHECK17-NEXT: [[TMP5:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 82003// CHECK17-NEXT: [[TMP6:%.*]] = load ptr, ptr [[A_ADDR]], align 82004// CHECK17-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 8, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPg_l152.omp_outlined, ptr [[TMP4]], ptr [[N_ADDR]], i64 [[TMP0]], ptr [[TMP5]], ptr [[TMP6]], i64 [[TMP1]], i64 [[TMP2]], ptr [[TMP3]])2005// CHECK17-NEXT: ret void2006//2007//2008// CHECK17-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPg_l152.omp_outlined2009// CHECK17-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], ptr [[S:%.*]], ptr nonnull align 4 dereferenceable(4) [[N:%.*]], i64 [[VLA:%.*]], ptr [[VLA1:%.*]], ptr [[A:%.*]], i64 [[VLA2:%.*]], i64 [[VLA4:%.*]], ptr nonnull align 8 dereferenceable(8) [[VLA26:%.*]]) #[[ATTR2]] {2010// CHECK17-NEXT: entry:2011// CHECK17-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 82012// CHECK17-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 82013// CHECK17-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 82014// CHECK17-NEXT: [[N_ADDR:%.*]] = alloca ptr, align 82015// CHECK17-NEXT: [[VLA_ADDR:%.*]] = alloca i64, align 82016// CHECK17-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 82017// CHECK17-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 82018// CHECK17-NEXT: [[VLA_ADDR3:%.*]] = alloca i64, align 82019// CHECK17-NEXT: [[VLA_ADDR5:%.*]] = alloca i64, align 82020// CHECK17-NEXT: [[VLA2_ADDR:%.*]] = alloca ptr, align 82021// CHECK17-NEXT: [[SAVED_STACK:%.*]] = alloca ptr, align 82022// CHECK17-NEXT: [[__VLA_EXPR0:%.*]] = alloca i64, align 82023// CHECK17-NEXT: [[__VLA_EXPR1:%.*]] = alloca i64, align 82024// CHECK17-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 82025// CHECK17-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 82026// CHECK17-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 82027// CHECK17-NEXT: store ptr [[N]], ptr [[N_ADDR]], align 82028// CHECK17-NEXT: store i64 [[VLA]], ptr [[VLA_ADDR]], align 82029// CHECK17-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 82030// CHECK17-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 82031// CHECK17-NEXT: store i64 [[VLA2]], ptr [[VLA_ADDR3]], align 82032// CHECK17-NEXT: store i64 [[VLA4]], ptr [[VLA_ADDR5]], align 82033// CHECK17-NEXT: store ptr [[VLA26]], ptr [[VLA2_ADDR]], align 82034// CHECK17-NEXT: [[TMP0:%.*]] = load ptr, ptr [[N_ADDR]], align 8, !nonnull [[META8]], !align [[META10:![0-9]+]]2035// CHECK17-NEXT: [[TMP1:%.*]] = load i64, ptr [[VLA_ADDR]], align 82036// CHECK17-NEXT: [[TMP2:%.*]] = load i64, ptr [[VLA_ADDR3]], align 82037// CHECK17-NEXT: [[TMP3:%.*]] = load i64, ptr [[VLA_ADDR5]], align 82038// CHECK17-NEXT: [[TMP4:%.*]] = load ptr, ptr [[VLA2_ADDR]], align 8, !nonnull [[META8]], !align [[META9]]2039// CHECK17-NEXT: [[TMP5:%.*]] = call ptr @llvm.stacksave.p0()2040// CHECK17-NEXT: store ptr [[TMP5]], ptr [[SAVED_STACK]], align 82041// CHECK17-NEXT: [[TMP6:%.*]] = mul nuw i64 [[TMP2]], [[TMP3]]2042// CHECK17-NEXT: [[VLA7:%.*]] = alloca double, i64 [[TMP6]], align 1282043// CHECK17-NEXT: store i64 [[TMP2]], ptr [[__VLA_EXPR0]], align 82044// CHECK17-NEXT: store i64 [[TMP3]], ptr [[__VLA_EXPR1]], align 82045// CHECK17-NEXT: [[TMP7:%.*]] = mul nuw i64 [[TMP2]], [[TMP3]]2046// CHECK17-NEXT: [[TMP8:%.*]] = mul nuw i64 [[TMP7]], 82047// CHECK17-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 128 [[VLA7]], ptr align 128 [[TMP4]], i64 [[TMP8]], i1 false)2048// CHECK17-NEXT: [[TMP9:%.*]] = load ptr, ptr [[S_ADDR]], align 82049// CHECK17-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [[STRUCT_ST:%.*]], ptr [[TMP9]], i64 02050// CHECK17-NEXT: [[TMP10:%.*]] = load ptr, ptr [[S_ADDR]], align 82051// CHECK17-NEXT: [[TMP11:%.*]] = load i32, ptr [[TMP0]], align 42052// CHECK17-NEXT: [[TMP12:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 82053// CHECK17-NEXT: call void @_ZN2St7St_funcEPS_iPg(ptr nonnull align 4 dereferenceable(8) [[ARRAYIDX]], ptr [[TMP10]], i32 signext [[TMP11]], ptr [[TMP12]])2054// CHECK17-NEXT: [[TMP13:%.*]] = load ptr, ptr [[SAVED_STACK]], align 82055// CHECK17-NEXT: call void @llvm.stackrestore.p0(ptr [[TMP13]])2056// CHECK17-NEXT: ret void2057//2058//2059// CHECK17-LABEL: define {{[^@]+}}@_ZN2St7St_funcEPS_iPg2060// CHECK17-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]], ptr [[S:%.*]], i32 signext [[N:%.*]], ptr [[VLA1:%.*]]) #[[ATTR0]] comdat {2061// CHECK17-NEXT: entry:2062// CHECK17-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 82063// CHECK17-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 82064// CHECK17-NEXT: [[N_ADDR:%.*]] = alloca i32, align 42065// CHECK17-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 82066// CHECK17-NEXT: [[SAVED_STACK:%.*]] = alloca ptr, align 82067// CHECK17-NEXT: [[__VLA_EXPR0:%.*]] = alloca i64, align 82068// CHECK17-NEXT: [[__VLA_EXPR1:%.*]] = alloca i64, align 82069// CHECK17-NEXT: [[N_CASTED:%.*]] = alloca i64, align 82070// CHECK17-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [10 x ptr], align 82071// CHECK17-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [10 x ptr], align 82072// CHECK17-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [10 x ptr], align 82073// CHECK17-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [10 x i64], align 82074// CHECK17-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 82075// CHECK17-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 82076// CHECK17-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 82077// CHECK17-NEXT: store i32 [[N]], ptr [[N_ADDR]], align 42078// CHECK17-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 82079// CHECK17-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 82080// CHECK17-NEXT: [[TMP0:%.*]] = load i32, ptr [[N_ADDR]], align 42081// CHECK17-NEXT: [[TMP1:%.*]] = zext i32 [[TMP0]] to i642082// CHECK17-NEXT: [[TMP2:%.*]] = load i32, ptr [[N_ADDR]], align 42083// CHECK17-NEXT: [[TMP3:%.*]] = zext i32 [[TMP2]] to i642084// CHECK17-NEXT: [[TMP4:%.*]] = load i32, ptr [[N_ADDR]], align 42085// CHECK17-NEXT: [[TMP5:%.*]] = zext i32 [[TMP4]] to i642086// CHECK17-NEXT: [[TMP6:%.*]] = call ptr @llvm.stacksave.p0()2087// CHECK17-NEXT: store ptr [[TMP6]], ptr [[SAVED_STACK]], align 82088// CHECK17-NEXT: [[TMP7:%.*]] = mul nuw i64 [[TMP3]], [[TMP5]]2089// CHECK17-NEXT: [[VLA:%.*]] = alloca double, i64 [[TMP7]], align 1282090// CHECK17-NEXT: store i64 [[TMP3]], ptr [[__VLA_EXPR0]], align 82091// CHECK17-NEXT: store i64 [[TMP5]], ptr [[__VLA_EXPR1]], align 82092// CHECK17-NEXT: [[B:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[THIS1]], i32 0, i32 12093// CHECK17-NEXT: [[TMP8:%.*]] = load i32, ptr [[B]], align 42094// CHECK17-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[THIS1]], i32 0, i32 02095// CHECK17-NEXT: store i32 [[TMP8]], ptr [[A]], align 42096// CHECK17-NEXT: [[TMP9:%.*]] = load ptr, ptr [[S_ADDR]], align 82097// CHECK17-NEXT: [[TMP10:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 82098// CHECK17-NEXT: [[TMP11:%.*]] = load i32, ptr [[N_ADDR]], align 42099// CHECK17-NEXT: store i32 [[TMP11]], ptr [[N_CASTED]], align 42100// CHECK17-NEXT: [[TMP12:%.*]] = load i64, ptr [[N_CASTED]], align 82101// CHECK17-NEXT: [[TMP13:%.*]] = mul nuw i64 [[TMP3]], [[TMP5]]2102// CHECK17-NEXT: [[TMP14:%.*]] = mul nuw i64 [[TMP13]], 82103// CHECK17-NEXT: [[B2:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[THIS1]], i32 0, i32 12104// CHECK17-NEXT: [[A3:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[THIS1]], i32 0, i32 02105// CHECK17-NEXT: [[TMP15:%.*]] = getelementptr i32, ptr [[B2]], i32 12106// CHECK17-NEXT: [[TMP16:%.*]] = ptrtoint ptr [[TMP15]] to i642107// CHECK17-NEXT: [[TMP17:%.*]] = ptrtoint ptr [[A3]] to i642108// CHECK17-NEXT: [[TMP18:%.*]] = sub i64 [[TMP16]], [[TMP17]]2109// CHECK17-NEXT: [[TMP19:%.*]] = sdiv exact i64 [[TMP18]], ptrtoint (ptr getelementptr (i8, ptr null, i32 1) to i64)2110// CHECK17-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 8 [[DOTOFFLOAD_SIZES]], ptr align 8 @.offload_sizes.1, i64 80, i1 false)2111// CHECK17-NEXT: [[TMP20:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 02112// CHECK17-NEXT: store ptr [[TMP9]], ptr [[TMP20]], align 82113// CHECK17-NEXT: [[TMP21:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 02114// CHECK17-NEXT: store ptr [[TMP9]], ptr [[TMP21]], align 82115// CHECK17-NEXT: [[TMP22:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 02116// CHECK17-NEXT: store ptr null, ptr [[TMP22]], align 82117// CHECK17-NEXT: [[TMP23:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 12118// CHECK17-NEXT: store i64 [[TMP1]], ptr [[TMP23]], align 82119// CHECK17-NEXT: [[TMP24:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 12120// CHECK17-NEXT: store i64 [[TMP1]], ptr [[TMP24]], align 82121// CHECK17-NEXT: [[TMP25:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 12122// CHECK17-NEXT: store ptr null, ptr [[TMP25]], align 82123// CHECK17-NEXT: [[TMP26:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 22124// CHECK17-NEXT: store ptr [[TMP10]], ptr [[TMP26]], align 82125// CHECK17-NEXT: [[TMP27:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 22126// CHECK17-NEXT: store ptr [[TMP10]], ptr [[TMP27]], align 82127// CHECK17-NEXT: [[TMP28:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 22128// CHECK17-NEXT: store ptr null, ptr [[TMP28]], align 82129// CHECK17-NEXT: [[TMP29:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 32130// CHECK17-NEXT: store i64 [[TMP3]], ptr [[TMP29]], align 82131// CHECK17-NEXT: [[TMP30:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 32132// CHECK17-NEXT: store i64 [[TMP3]], ptr [[TMP30]], align 82133// CHECK17-NEXT: [[TMP31:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 32134// CHECK17-NEXT: store ptr null, ptr [[TMP31]], align 82135// CHECK17-NEXT: [[TMP32:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 42136// CHECK17-NEXT: store i64 [[TMP5]], ptr [[TMP32]], align 82137// CHECK17-NEXT: [[TMP33:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 42138// CHECK17-NEXT: store i64 [[TMP5]], ptr [[TMP33]], align 82139// CHECK17-NEXT: [[TMP34:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 42140// CHECK17-NEXT: store ptr null, ptr [[TMP34]], align 82141// CHECK17-NEXT: [[TMP35:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 52142// CHECK17-NEXT: store ptr [[VLA]], ptr [[TMP35]], align 82143// CHECK17-NEXT: [[TMP36:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 52144// CHECK17-NEXT: store ptr [[VLA]], ptr [[TMP36]], align 82145// CHECK17-NEXT: [[TMP37:%.*]] = getelementptr inbounds [10 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 52146// CHECK17-NEXT: store i64 [[TMP14]], ptr [[TMP37]], align 82147// CHECK17-NEXT: [[TMP38:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 52148// CHECK17-NEXT: store ptr null, ptr [[TMP38]], align 82149// CHECK17-NEXT: [[TMP39:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 62150// CHECK17-NEXT: store ptr [[THIS1]], ptr [[TMP39]], align 82151// CHECK17-NEXT: [[TMP40:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 62152// CHECK17-NEXT: store ptr [[A3]], ptr [[TMP40]], align 82153// CHECK17-NEXT: [[TMP41:%.*]] = getelementptr inbounds [10 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 62154// CHECK17-NEXT: store i64 [[TMP19]], ptr [[TMP41]], align 82155// CHECK17-NEXT: [[TMP42:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 62156// CHECK17-NEXT: store ptr null, ptr [[TMP42]], align 82157// CHECK17-NEXT: [[TMP43:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 72158// CHECK17-NEXT: store ptr [[THIS1]], ptr [[TMP43]], align 82159// CHECK17-NEXT: [[TMP44:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 72160// CHECK17-NEXT: store ptr [[B2]], ptr [[TMP44]], align 82161// CHECK17-NEXT: [[TMP45:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 72162// CHECK17-NEXT: store ptr null, ptr [[TMP45]], align 82163// CHECK17-NEXT: [[TMP46:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 82164// CHECK17-NEXT: store ptr [[THIS1]], ptr [[TMP46]], align 82165// CHECK17-NEXT: [[TMP47:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 82166// CHECK17-NEXT: store ptr [[A3]], ptr [[TMP47]], align 82167// CHECK17-NEXT: [[TMP48:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 82168// CHECK17-NEXT: store ptr null, ptr [[TMP48]], align 82169// CHECK17-NEXT: [[TMP49:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 92170// CHECK17-NEXT: store i64 [[TMP12]], ptr [[TMP49]], align 82171// CHECK17-NEXT: [[TMP50:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 92172// CHECK17-NEXT: store i64 [[TMP12]], ptr [[TMP50]], align 82173// CHECK17-NEXT: [[TMP51:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 92174// CHECK17-NEXT: store ptr null, ptr [[TMP51]], align 82175// CHECK17-NEXT: [[TMP52:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 02176// CHECK17-NEXT: [[TMP53:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 02177// CHECK17-NEXT: [[TMP54:%.*]] = getelementptr inbounds [10 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 02178// CHECK17-NEXT: [[TMP55:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 02179// CHECK17-NEXT: store i32 3, ptr [[TMP55]], align 42180// CHECK17-NEXT: [[TMP56:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12181// CHECK17-NEXT: store i32 10, ptr [[TMP56]], align 42182// CHECK17-NEXT: [[TMP57:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 22183// CHECK17-NEXT: store ptr [[TMP52]], ptr [[TMP57]], align 82184// CHECK17-NEXT: [[TMP58:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 32185// CHECK17-NEXT: store ptr [[TMP53]], ptr [[TMP58]], align 82186// CHECK17-NEXT: [[TMP59:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 42187// CHECK17-NEXT: store ptr [[TMP54]], ptr [[TMP59]], align 82188// CHECK17-NEXT: [[TMP60:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 52189// CHECK17-NEXT: store ptr @.offload_maptypes.2, ptr [[TMP60]], align 82190// CHECK17-NEXT: [[TMP61:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 62191// CHECK17-NEXT: store ptr null, ptr [[TMP61]], align 82192// CHECK17-NEXT: [[TMP62:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 72193// CHECK17-NEXT: store ptr null, ptr [[TMP62]], align 82194// CHECK17-NEXT: [[TMP63:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 82195// CHECK17-NEXT: store i64 0, ptr [[TMP63]], align 82196// CHECK17-NEXT: [[TMP64:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 92197// CHECK17-NEXT: store i64 0, ptr [[TMP64]], align 82198// CHECK17-NEXT: [[TMP65:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 102199// CHECK17-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP65]], align 42200// CHECK17-NEXT: [[TMP66:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 112201// CHECK17-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP66]], align 42202// CHECK17-NEXT: [[TMP67:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 122203// CHECK17-NEXT: store i32 0, ptr [[TMP67]], align 42204// CHECK17-NEXT: [[TMP68:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPg_l144.region_id, ptr [[KERNEL_ARGS]])2205// CHECK17-NEXT: [[TMP69:%.*]] = icmp ne i32 [[TMP68]], 02206// CHECK17-NEXT: br i1 [[TMP69]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]2207// CHECK17: omp_offload.failed:2208// CHECK17-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPg_l144(ptr [[TMP9]], i64 [[TMP1]], ptr [[TMP10]], i64 [[TMP3]], i64 [[TMP5]], ptr [[VLA]], ptr [[THIS1]], i64 [[TMP12]]) #[[ATTR4]]2209// CHECK17-NEXT: br label [[OMP_OFFLOAD_CONT]]2210// CHECK17: omp_offload.cont:2211// CHECK17-NEXT: [[TMP70:%.*]] = load ptr, ptr [[SAVED_STACK]], align 82212// CHECK17-NEXT: call void @llvm.stackrestore.p0(ptr [[TMP70]])2213// CHECK17-NEXT: ret void2214//2215//2216// CHECK17-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPg_l1442217// CHECK17-SAME: (ptr [[S:%.*]], i64 [[VLA:%.*]], ptr [[VLA1:%.*]], i64 [[VLA2:%.*]], i64 [[VLA4:%.*]], ptr nonnull align 8 dereferenceable(8) [[VLA26:%.*]], ptr [[THIS:%.*]], i64 [[N:%.*]]) #[[ATTR2]] {2218// CHECK17-NEXT: entry:2219// CHECK17-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 82220// CHECK17-NEXT: [[VLA_ADDR:%.*]] = alloca i64, align 82221// CHECK17-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 82222// CHECK17-NEXT: [[VLA_ADDR3:%.*]] = alloca i64, align 82223// CHECK17-NEXT: [[VLA_ADDR5:%.*]] = alloca i64, align 82224// CHECK17-NEXT: [[VLA2_ADDR:%.*]] = alloca ptr, align 82225// CHECK17-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 82226// CHECK17-NEXT: [[N_ADDR:%.*]] = alloca i64, align 82227// CHECK17-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 82228// CHECK17-NEXT: store i64 [[VLA]], ptr [[VLA_ADDR]], align 82229// CHECK17-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 82230// CHECK17-NEXT: store i64 [[VLA2]], ptr [[VLA_ADDR3]], align 82231// CHECK17-NEXT: store i64 [[VLA4]], ptr [[VLA_ADDR5]], align 82232// CHECK17-NEXT: store ptr [[VLA26]], ptr [[VLA2_ADDR]], align 82233// CHECK17-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 82234// CHECK17-NEXT: store i64 [[N]], ptr [[N_ADDR]], align 82235// CHECK17-NEXT: [[TMP0:%.*]] = load i64, ptr [[VLA_ADDR]], align 82236// CHECK17-NEXT: [[TMP1:%.*]] = load i64, ptr [[VLA_ADDR3]], align 82237// CHECK17-NEXT: [[TMP2:%.*]] = load i64, ptr [[VLA_ADDR5]], align 82238// CHECK17-NEXT: [[TMP3:%.*]] = load ptr, ptr [[VLA2_ADDR]], align 8, !nonnull [[META8]], !align [[META9]]2239// CHECK17-NEXT: [[TMP4:%.*]] = load ptr, ptr [[THIS_ADDR]], align 82240// CHECK17-NEXT: [[TMP5:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 82241// CHECK17-NEXT: [[TMP6:%.*]] = load ptr, ptr [[S_ADDR]], align 82242// CHECK17-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 8, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPg_l144.omp_outlined, i64 [[TMP0]], ptr [[TMP5]], ptr [[TMP4]], i64 [[TMP1]], i64 [[TMP2]], ptr [[TMP3]], ptr [[N_ADDR]], ptr [[TMP6]])2243// CHECK17-NEXT: ret void2244//2245//2246// CHECK17-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPg_l144.omp_outlined2247// CHECK17-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i64 [[VLA:%.*]], ptr [[VLA1:%.*]], ptr [[THIS:%.*]], i64 [[VLA2:%.*]], i64 [[VLA4:%.*]], ptr nonnull align 8 dereferenceable(8) [[VLA26:%.*]], ptr nonnull align 4 dereferenceable(4) [[N:%.*]], ptr [[S:%.*]]) #[[ATTR2]] {2248// CHECK17-NEXT: entry:2249// CHECK17-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 82250// CHECK17-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 82251// CHECK17-NEXT: [[VLA_ADDR:%.*]] = alloca i64, align 82252// CHECK17-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 82253// CHECK17-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 82254// CHECK17-NEXT: [[VLA_ADDR3:%.*]] = alloca i64, align 82255// CHECK17-NEXT: [[VLA_ADDR5:%.*]] = alloca i64, align 82256// CHECK17-NEXT: [[VLA2_ADDR:%.*]] = alloca ptr, align 82257// CHECK17-NEXT: [[N_ADDR:%.*]] = alloca ptr, align 82258// CHECK17-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 82259// CHECK17-NEXT: [[SAVED_STACK:%.*]] = alloca ptr, align 82260// CHECK17-NEXT: [[__VLA_EXPR0:%.*]] = alloca i64, align 82261// CHECK17-NEXT: [[__VLA_EXPR1:%.*]] = alloca i64, align 82262// CHECK17-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 82263// CHECK17-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 82264// CHECK17-NEXT: store i64 [[VLA]], ptr [[VLA_ADDR]], align 82265// CHECK17-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 82266// CHECK17-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 82267// CHECK17-NEXT: store i64 [[VLA2]], ptr [[VLA_ADDR3]], align 82268// CHECK17-NEXT: store i64 [[VLA4]], ptr [[VLA_ADDR5]], align 82269// CHECK17-NEXT: store ptr [[VLA26]], ptr [[VLA2_ADDR]], align 82270// CHECK17-NEXT: store ptr [[N]], ptr [[N_ADDR]], align 82271// CHECK17-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 82272// CHECK17-NEXT: [[TMP0:%.*]] = load i64, ptr [[VLA_ADDR]], align 82273// CHECK17-NEXT: [[TMP1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 82274// CHECK17-NEXT: [[TMP2:%.*]] = load i64, ptr [[VLA_ADDR3]], align 82275// CHECK17-NEXT: [[TMP3:%.*]] = load i64, ptr [[VLA_ADDR5]], align 82276// CHECK17-NEXT: [[TMP4:%.*]] = load ptr, ptr [[VLA2_ADDR]], align 8, !nonnull [[META8]], !align [[META9]]2277// CHECK17-NEXT: [[TMP5:%.*]] = load ptr, ptr [[N_ADDR]], align 8, !nonnull [[META8]], !align [[META10]]2278// CHECK17-NEXT: [[TMP6:%.*]] = call ptr @llvm.stacksave.p0()2279// CHECK17-NEXT: store ptr [[TMP6]], ptr [[SAVED_STACK]], align 82280// CHECK17-NEXT: [[TMP7:%.*]] = mul nuw i64 [[TMP2]], [[TMP3]]2281// CHECK17-NEXT: [[VLA7:%.*]] = alloca double, i64 [[TMP7]], align 1282282// CHECK17-NEXT: store i64 [[TMP2]], ptr [[__VLA_EXPR0]], align 82283// CHECK17-NEXT: store i64 [[TMP3]], ptr [[__VLA_EXPR1]], align 82284// CHECK17-NEXT: [[TMP8:%.*]] = mul nuw i64 [[TMP2]], [[TMP3]]2285// CHECK17-NEXT: [[TMP9:%.*]] = mul nuw i64 [[TMP8]], 82286// CHECK17-NEXT: call void @llvm.memcpy.p0.p0.i64(ptr align 128 [[VLA7]], ptr align 128 [[TMP4]], i64 [[TMP9]], i1 false)2287// CHECK17-NEXT: [[B:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[TMP1]], i32 0, i32 12288// CHECK17-NEXT: [[TMP10:%.*]] = load i32, ptr [[B]], align 42289// CHECK17-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[TMP1]], i32 0, i32 02290// CHECK17-NEXT: store i32 [[TMP10]], ptr [[A]], align 42291// CHECK17-NEXT: [[CONV:%.*]] = sitofp i32 [[TMP10]] to double2292// CHECK17-NEXT: [[TMP11:%.*]] = mul nsw i64 1, [[TMP3]]2293// CHECK17-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds double, ptr [[VLA7]], i64 [[TMP11]]2294// CHECK17-NEXT: [[TMP12:%.*]] = load i32, ptr [[TMP5]], align 42295// CHECK17-NEXT: [[SUB:%.*]] = sub nsw i32 [[TMP12]], 12296// CHECK17-NEXT: [[IDXPROM:%.*]] = sext i32 [[SUB]] to i642297// CHECK17-NEXT: [[ARRAYIDX8:%.*]] = getelementptr inbounds double, ptr [[ARRAYIDX]], i64 [[IDXPROM]]2298// CHECK17-NEXT: store double [[CONV]], ptr [[ARRAYIDX8]], align 82299// CHECK17-NEXT: [[CONV9:%.*]] = fpext double [[CONV]] to ppc_fp1282300// CHECK17-NEXT: [[TMP13:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 82301// CHECK17-NEXT: [[B10:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[TMP1]], i32 0, i32 12302// CHECK17-NEXT: [[TMP14:%.*]] = load i32, ptr [[B10]], align 42303// CHECK17-NEXT: [[IDXPROM11:%.*]] = sext i32 [[TMP14]] to i642304// CHECK17-NEXT: [[ARRAYIDX12:%.*]] = getelementptr inbounds ppc_fp128, ptr [[TMP13]], i64 [[IDXPROM11]]2305// CHECK17-NEXT: store ppc_fp128 [[CONV9]], ptr [[ARRAYIDX12]], align 162306// CHECK17-NEXT: [[TMP15:%.*]] = load ptr, ptr [[SAVED_STACK]], align 82307// CHECK17-NEXT: call void @llvm.stackrestore.p0(ptr [[TMP15]])2308// CHECK17-NEXT: ret void2309//2310//2311// CHECK19-LABEL: define {{[^@]+}}@_Z10array_funcPfP2StiPe2312// CHECK19-SAME: (ptr [[A:%.*]], ptr [[S:%.*]], i32 [[N:%.*]], ptr [[VLA1:%.*]]) #[[ATTR0:[0-9]+]] {2313// CHECK19-NEXT: entry:2314// CHECK19-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 42315// CHECK19-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 42316// CHECK19-NEXT: [[N_ADDR:%.*]] = alloca i32, align 42317// CHECK19-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 42318// CHECK19-NEXT: [[SAVED_STACK:%.*]] = alloca ptr, align 42319// CHECK19-NEXT: [[__VLA_EXPR0:%.*]] = alloca i32, align 42320// CHECK19-NEXT: [[__VLA_EXPR1:%.*]] = alloca i32, align 42321// CHECK19-NEXT: [[N_CASTED:%.*]] = alloca i32, align 42322// CHECK19-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [8 x ptr], align 42323// CHECK19-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [8 x ptr], align 42324// CHECK19-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [8 x ptr], align 42325// CHECK19-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [8 x i64], align 42326// CHECK19-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 82327// CHECK19-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 42328// CHECK19-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 42329// CHECK19-NEXT: store i32 [[N]], ptr [[N_ADDR]], align 42330// CHECK19-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 42331// CHECK19-NEXT: [[TMP0:%.*]] = load i32, ptr [[N_ADDR]], align 42332// CHECK19-NEXT: [[TMP1:%.*]] = load i32, ptr [[N_ADDR]], align 42333// CHECK19-NEXT: [[TMP2:%.*]] = load i32, ptr [[N_ADDR]], align 42334// CHECK19-NEXT: [[TMP3:%.*]] = call ptr @llvm.stacksave.p0()2335// CHECK19-NEXT: store ptr [[TMP3]], ptr [[SAVED_STACK]], align 42336// CHECK19-NEXT: [[TMP4:%.*]] = mul nuw i32 [[TMP1]], [[TMP2]]2337// CHECK19-NEXT: [[VLA:%.*]] = alloca double, i32 [[TMP4]], align 1282338// CHECK19-NEXT: store i32 [[TMP1]], ptr [[__VLA_EXPR0]], align 42339// CHECK19-NEXT: store i32 [[TMP2]], ptr [[__VLA_EXPR1]], align 42340// CHECK19-NEXT: [[TMP5:%.*]] = load ptr, ptr [[A_ADDR]], align 42341// CHECK19-NEXT: [[TMP6:%.*]] = load ptr, ptr [[S_ADDR]], align 42342// CHECK19-NEXT: [[TMP7:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 42343// CHECK19-NEXT: [[TMP8:%.*]] = load i32, ptr [[N_ADDR]], align 42344// CHECK19-NEXT: store i32 [[TMP8]], ptr [[N_CASTED]], align 42345// CHECK19-NEXT: [[TMP9:%.*]] = load i32, ptr [[N_CASTED]], align 42346// CHECK19-NEXT: [[TMP10:%.*]] = mul nuw i32 [[TMP1]], [[TMP2]]2347// CHECK19-NEXT: [[TMP11:%.*]] = mul nuw i32 [[TMP10]], 82348// CHECK19-NEXT: [[TMP12:%.*]] = sext i32 [[TMP11]] to i642349// CHECK19-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[DOTOFFLOAD_SIZES]], ptr align 4 @.offload_sizes, i32 64, i1 false)2350// CHECK19-NEXT: [[TMP13:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 02351// CHECK19-NEXT: store ptr [[TMP5]], ptr [[TMP13]], align 42352// CHECK19-NEXT: [[TMP14:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 02353// CHECK19-NEXT: store ptr [[TMP5]], ptr [[TMP14]], align 42354// CHECK19-NEXT: [[TMP15:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 02355// CHECK19-NEXT: store ptr null, ptr [[TMP15]], align 42356// CHECK19-NEXT: [[TMP16:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 12357// CHECK19-NEXT: store ptr [[TMP6]], ptr [[TMP16]], align 42358// CHECK19-NEXT: [[TMP17:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 12359// CHECK19-NEXT: store ptr [[TMP6]], ptr [[TMP17]], align 42360// CHECK19-NEXT: [[TMP18:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 12361// CHECK19-NEXT: store ptr null, ptr [[TMP18]], align 42362// CHECK19-NEXT: [[TMP19:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 22363// CHECK19-NEXT: store i32 [[TMP0]], ptr [[TMP19]], align 42364// CHECK19-NEXT: [[TMP20:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 22365// CHECK19-NEXT: store i32 [[TMP0]], ptr [[TMP20]], align 42366// CHECK19-NEXT: [[TMP21:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 22367// CHECK19-NEXT: store ptr null, ptr [[TMP21]], align 42368// CHECK19-NEXT: [[TMP22:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 32369// CHECK19-NEXT: store ptr [[TMP7]], ptr [[TMP22]], align 42370// CHECK19-NEXT: [[TMP23:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 32371// CHECK19-NEXT: store ptr [[TMP7]], ptr [[TMP23]], align 42372// CHECK19-NEXT: [[TMP24:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 32373// CHECK19-NEXT: store ptr null, ptr [[TMP24]], align 42374// CHECK19-NEXT: [[TMP25:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 42375// CHECK19-NEXT: store i32 [[TMP1]], ptr [[TMP25]], align 42376// CHECK19-NEXT: [[TMP26:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 42377// CHECK19-NEXT: store i32 [[TMP1]], ptr [[TMP26]], align 42378// CHECK19-NEXT: [[TMP27:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 42379// CHECK19-NEXT: store ptr null, ptr [[TMP27]], align 42380// CHECK19-NEXT: [[TMP28:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 52381// CHECK19-NEXT: store i32 [[TMP2]], ptr [[TMP28]], align 42382// CHECK19-NEXT: [[TMP29:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 52383// CHECK19-NEXT: store i32 [[TMP2]], ptr [[TMP29]], align 42384// CHECK19-NEXT: [[TMP30:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 52385// CHECK19-NEXT: store ptr null, ptr [[TMP30]], align 42386// CHECK19-NEXT: [[TMP31:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 62387// CHECK19-NEXT: store ptr [[VLA]], ptr [[TMP31]], align 42388// CHECK19-NEXT: [[TMP32:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 62389// CHECK19-NEXT: store ptr [[VLA]], ptr [[TMP32]], align 42390// CHECK19-NEXT: [[TMP33:%.*]] = getelementptr inbounds [8 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 62391// CHECK19-NEXT: store i64 [[TMP12]], ptr [[TMP33]], align 42392// CHECK19-NEXT: [[TMP34:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 62393// CHECK19-NEXT: store ptr null, ptr [[TMP34]], align 42394// CHECK19-NEXT: [[TMP35:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 72395// CHECK19-NEXT: store i32 [[TMP9]], ptr [[TMP35]], align 42396// CHECK19-NEXT: [[TMP36:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 72397// CHECK19-NEXT: store i32 [[TMP9]], ptr [[TMP36]], align 42398// CHECK19-NEXT: [[TMP37:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 72399// CHECK19-NEXT: store ptr null, ptr [[TMP37]], align 42400// CHECK19-NEXT: [[TMP38:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 02401// CHECK19-NEXT: [[TMP39:%.*]] = getelementptr inbounds [8 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 02402// CHECK19-NEXT: [[TMP40:%.*]] = getelementptr inbounds [8 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 02403// CHECK19-NEXT: [[TMP41:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 02404// CHECK19-NEXT: store i32 3, ptr [[TMP41]], align 42405// CHECK19-NEXT: [[TMP42:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12406// CHECK19-NEXT: store i32 8, ptr [[TMP42]], align 42407// CHECK19-NEXT: [[TMP43:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 22408// CHECK19-NEXT: store ptr [[TMP38]], ptr [[TMP43]], align 42409// CHECK19-NEXT: [[TMP44:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 32410// CHECK19-NEXT: store ptr [[TMP39]], ptr [[TMP44]], align 42411// CHECK19-NEXT: [[TMP45:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 42412// CHECK19-NEXT: store ptr [[TMP40]], ptr [[TMP45]], align 42413// CHECK19-NEXT: [[TMP46:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 52414// CHECK19-NEXT: store ptr @.offload_maptypes, ptr [[TMP46]], align 42415// CHECK19-NEXT: [[TMP47:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 62416// CHECK19-NEXT: store ptr null, ptr [[TMP47]], align 42417// CHECK19-NEXT: [[TMP48:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 72418// CHECK19-NEXT: store ptr null, ptr [[TMP48]], align 42419// CHECK19-NEXT: [[TMP49:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 82420// CHECK19-NEXT: store i64 0, ptr [[TMP49]], align 82421// CHECK19-NEXT: [[TMP50:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 92422// CHECK19-NEXT: store i64 0, ptr [[TMP50]], align 82423// CHECK19-NEXT: [[TMP51:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 102424// CHECK19-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP51]], align 42425// CHECK19-NEXT: [[TMP52:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 112426// CHECK19-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP52]], align 42427// CHECK19-NEXT: [[TMP53:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 122428// CHECK19-NEXT: store i32 0, ptr [[TMP53]], align 42429// CHECK19-NEXT: [[TMP54:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPe_l152.region_id, ptr [[KERNEL_ARGS]])2430// CHECK19-NEXT: [[TMP55:%.*]] = icmp ne i32 [[TMP54]], 02431// CHECK19-NEXT: br i1 [[TMP55]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]2432// CHECK19: omp_offload.failed:2433// CHECK19-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPe_l152(ptr [[TMP5]], ptr [[TMP6]], i32 [[TMP0]], ptr [[TMP7]], i32 [[TMP1]], i32 [[TMP2]], ptr [[VLA]], i32 [[TMP9]]) #[[ATTR4:[0-9]+]]2434// CHECK19-NEXT: br label [[OMP_OFFLOAD_CONT]]2435// CHECK19: omp_offload.cont:2436// CHECK19-NEXT: [[TMP56:%.*]] = load ptr, ptr [[SAVED_STACK]], align 42437// CHECK19-NEXT: call void @llvm.stackrestore.p0(ptr [[TMP56]])2438// CHECK19-NEXT: ret void2439//2440//2441// CHECK19-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPe_l1522442// CHECK19-SAME: (ptr [[A:%.*]], ptr [[S:%.*]], i32 [[VLA:%.*]], ptr [[VLA1:%.*]], i32 [[VLA2:%.*]], i32 [[VLA4:%.*]], ptr nonnull align 4 dereferenceable(8) [[VLA26:%.*]], i32 [[N:%.*]]) #[[ATTR2:[0-9]+]] {2443// CHECK19-NEXT: entry:2444// CHECK19-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 42445// CHECK19-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 42446// CHECK19-NEXT: [[VLA_ADDR:%.*]] = alloca i32, align 42447// CHECK19-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 42448// CHECK19-NEXT: [[VLA_ADDR3:%.*]] = alloca i32, align 42449// CHECK19-NEXT: [[VLA_ADDR5:%.*]] = alloca i32, align 42450// CHECK19-NEXT: [[VLA2_ADDR:%.*]] = alloca ptr, align 42451// CHECK19-NEXT: [[N_ADDR:%.*]] = alloca i32, align 42452// CHECK19-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 42453// CHECK19-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 42454// CHECK19-NEXT: store i32 [[VLA]], ptr [[VLA_ADDR]], align 42455// CHECK19-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 42456// CHECK19-NEXT: store i32 [[VLA2]], ptr [[VLA_ADDR3]], align 42457// CHECK19-NEXT: store i32 [[VLA4]], ptr [[VLA_ADDR5]], align 42458// CHECK19-NEXT: store ptr [[VLA26]], ptr [[VLA2_ADDR]], align 42459// CHECK19-NEXT: store i32 [[N]], ptr [[N_ADDR]], align 42460// CHECK19-NEXT: [[TMP0:%.*]] = load i32, ptr [[VLA_ADDR]], align 42461// CHECK19-NEXT: [[TMP1:%.*]] = load i32, ptr [[VLA_ADDR3]], align 42462// CHECK19-NEXT: [[TMP2:%.*]] = load i32, ptr [[VLA_ADDR5]], align 42463// CHECK19-NEXT: [[TMP3:%.*]] = load ptr, ptr [[VLA2_ADDR]], align 4, !nonnull [[META8:![0-9]+]], !align [[META9:![0-9]+]]2464// CHECK19-NEXT: [[TMP4:%.*]] = load ptr, ptr [[S_ADDR]], align 42465// CHECK19-NEXT: [[TMP5:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 42466// CHECK19-NEXT: [[TMP6:%.*]] = load ptr, ptr [[A_ADDR]], align 42467// CHECK19-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 8, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPe_l152.omp_outlined, ptr [[TMP4]], ptr [[N_ADDR]], i32 [[TMP0]], ptr [[TMP5]], ptr [[TMP6]], i32 [[TMP1]], i32 [[TMP2]], ptr [[TMP3]])2468// CHECK19-NEXT: ret void2469//2470//2471// CHECK19-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z10array_funcPfP2StiPe_l152.omp_outlined2472// CHECK19-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], ptr [[S:%.*]], ptr nonnull align 4 dereferenceable(4) [[N:%.*]], i32 [[VLA:%.*]], ptr [[VLA1:%.*]], ptr [[A:%.*]], i32 [[VLA2:%.*]], i32 [[VLA4:%.*]], ptr nonnull align 4 dereferenceable(8) [[VLA26:%.*]]) #[[ATTR2]] {2473// CHECK19-NEXT: entry:2474// CHECK19-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 42475// CHECK19-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 42476// CHECK19-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 42477// CHECK19-NEXT: [[N_ADDR:%.*]] = alloca ptr, align 42478// CHECK19-NEXT: [[VLA_ADDR:%.*]] = alloca i32, align 42479// CHECK19-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 42480// CHECK19-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 42481// CHECK19-NEXT: [[VLA_ADDR3:%.*]] = alloca i32, align 42482// CHECK19-NEXT: [[VLA_ADDR5:%.*]] = alloca i32, align 42483// CHECK19-NEXT: [[VLA2_ADDR:%.*]] = alloca ptr, align 42484// CHECK19-NEXT: [[SAVED_STACK:%.*]] = alloca ptr, align 42485// CHECK19-NEXT: [[__VLA_EXPR0:%.*]] = alloca i32, align 42486// CHECK19-NEXT: [[__VLA_EXPR1:%.*]] = alloca i32, align 42487// CHECK19-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 42488// CHECK19-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 42489// CHECK19-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 42490// CHECK19-NEXT: store ptr [[N]], ptr [[N_ADDR]], align 42491// CHECK19-NEXT: store i32 [[VLA]], ptr [[VLA_ADDR]], align 42492// CHECK19-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 42493// CHECK19-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 42494// CHECK19-NEXT: store i32 [[VLA2]], ptr [[VLA_ADDR3]], align 42495// CHECK19-NEXT: store i32 [[VLA4]], ptr [[VLA_ADDR5]], align 42496// CHECK19-NEXT: store ptr [[VLA26]], ptr [[VLA2_ADDR]], align 42497// CHECK19-NEXT: [[TMP0:%.*]] = load ptr, ptr [[N_ADDR]], align 4, !nonnull [[META8]], !align [[META9]]2498// CHECK19-NEXT: [[TMP1:%.*]] = load i32, ptr [[VLA_ADDR]], align 42499// CHECK19-NEXT: [[TMP2:%.*]] = load i32, ptr [[VLA_ADDR3]], align 42500// CHECK19-NEXT: [[TMP3:%.*]] = load i32, ptr [[VLA_ADDR5]], align 42501// CHECK19-NEXT: [[TMP4:%.*]] = load ptr, ptr [[VLA2_ADDR]], align 4, !nonnull [[META8]], !align [[META9]]2502// CHECK19-NEXT: [[TMP5:%.*]] = call ptr @llvm.stacksave.p0()2503// CHECK19-NEXT: store ptr [[TMP5]], ptr [[SAVED_STACK]], align 42504// CHECK19-NEXT: [[TMP6:%.*]] = mul nuw i32 [[TMP2]], [[TMP3]]2505// CHECK19-NEXT: [[VLA7:%.*]] = alloca double, i32 [[TMP6]], align 1282506// CHECK19-NEXT: store i32 [[TMP2]], ptr [[__VLA_EXPR0]], align 42507// CHECK19-NEXT: store i32 [[TMP3]], ptr [[__VLA_EXPR1]], align 42508// CHECK19-NEXT: [[TMP7:%.*]] = mul nuw i32 [[TMP2]], [[TMP3]]2509// CHECK19-NEXT: [[TMP8:%.*]] = mul nuw i32 [[TMP7]], 82510// CHECK19-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 128 [[VLA7]], ptr align 128 [[TMP4]], i32 [[TMP8]], i1 false)2511// CHECK19-NEXT: [[TMP9:%.*]] = load ptr, ptr [[S_ADDR]], align 42512// CHECK19-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [[STRUCT_ST:%.*]], ptr [[TMP9]], i32 02513// CHECK19-NEXT: [[TMP10:%.*]] = load ptr, ptr [[S_ADDR]], align 42514// CHECK19-NEXT: [[TMP11:%.*]] = load i32, ptr [[TMP0]], align 42515// CHECK19-NEXT: [[TMP12:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 42516// CHECK19-NEXT: call void @_ZN2St7St_funcEPS_iPe(ptr nonnull align 4 dereferenceable(8) [[ARRAYIDX]], ptr [[TMP10]], i32 [[TMP11]], ptr [[TMP12]])2517// CHECK19-NEXT: [[TMP13:%.*]] = load ptr, ptr [[SAVED_STACK]], align 42518// CHECK19-NEXT: call void @llvm.stackrestore.p0(ptr [[TMP13]])2519// CHECK19-NEXT: ret void2520//2521//2522// CHECK19-LABEL: define {{[^@]+}}@_ZN2St7St_funcEPS_iPe2523// CHECK19-SAME: (ptr nonnull align 4 dereferenceable(8) [[THIS:%.*]], ptr [[S:%.*]], i32 [[N:%.*]], ptr [[VLA1:%.*]]) #[[ATTR0]] comdat align 2 {2524// CHECK19-NEXT: entry:2525// CHECK19-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 42526// CHECK19-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 42527// CHECK19-NEXT: [[N_ADDR:%.*]] = alloca i32, align 42528// CHECK19-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 42529// CHECK19-NEXT: [[SAVED_STACK:%.*]] = alloca ptr, align 42530// CHECK19-NEXT: [[__VLA_EXPR0:%.*]] = alloca i32, align 42531// CHECK19-NEXT: [[__VLA_EXPR1:%.*]] = alloca i32, align 42532// CHECK19-NEXT: [[N_CASTED:%.*]] = alloca i32, align 42533// CHECK19-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [10 x ptr], align 42534// CHECK19-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [10 x ptr], align 42535// CHECK19-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [10 x ptr], align 42536// CHECK19-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [10 x i64], align 42537// CHECK19-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 82538// CHECK19-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 42539// CHECK19-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 42540// CHECK19-NEXT: store i32 [[N]], ptr [[N_ADDR]], align 42541// CHECK19-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 42542// CHECK19-NEXT: [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 42543// CHECK19-NEXT: [[TMP0:%.*]] = load i32, ptr [[N_ADDR]], align 42544// CHECK19-NEXT: [[TMP1:%.*]] = load i32, ptr [[N_ADDR]], align 42545// CHECK19-NEXT: [[TMP2:%.*]] = load i32, ptr [[N_ADDR]], align 42546// CHECK19-NEXT: [[TMP3:%.*]] = call ptr @llvm.stacksave.p0()2547// CHECK19-NEXT: store ptr [[TMP3]], ptr [[SAVED_STACK]], align 42548// CHECK19-NEXT: [[TMP4:%.*]] = mul nuw i32 [[TMP1]], [[TMP2]]2549// CHECK19-NEXT: [[VLA:%.*]] = alloca double, i32 [[TMP4]], align 1282550// CHECK19-NEXT: store i32 [[TMP1]], ptr [[__VLA_EXPR0]], align 42551// CHECK19-NEXT: store i32 [[TMP2]], ptr [[__VLA_EXPR1]], align 42552// CHECK19-NEXT: [[B:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[THIS1]], i32 0, i32 12553// CHECK19-NEXT: [[TMP5:%.*]] = load i32, ptr [[B]], align 42554// CHECK19-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[THIS1]], i32 0, i32 02555// CHECK19-NEXT: store i32 [[TMP5]], ptr [[A]], align 42556// CHECK19-NEXT: [[TMP6:%.*]] = load ptr, ptr [[S_ADDR]], align 42557// CHECK19-NEXT: [[TMP7:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 42558// CHECK19-NEXT: [[TMP8:%.*]] = load i32, ptr [[N_ADDR]], align 42559// CHECK19-NEXT: store i32 [[TMP8]], ptr [[N_CASTED]], align 42560// CHECK19-NEXT: [[TMP9:%.*]] = load i32, ptr [[N_CASTED]], align 42561// CHECK19-NEXT: [[TMP10:%.*]] = mul nuw i32 [[TMP1]], [[TMP2]]2562// CHECK19-NEXT: [[TMP11:%.*]] = mul nuw i32 [[TMP10]], 82563// CHECK19-NEXT: [[TMP12:%.*]] = sext i32 [[TMP11]] to i642564// CHECK19-NEXT: [[B2:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[THIS1]], i32 0, i32 12565// CHECK19-NEXT: [[A3:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[THIS1]], i32 0, i32 02566// CHECK19-NEXT: [[TMP13:%.*]] = getelementptr i32, ptr [[B2]], i32 12567// CHECK19-NEXT: [[TMP14:%.*]] = ptrtoint ptr [[TMP13]] to i642568// CHECK19-NEXT: [[TMP15:%.*]] = ptrtoint ptr [[A3]] to i642569// CHECK19-NEXT: [[TMP16:%.*]] = sub i64 [[TMP14]], [[TMP15]]2570// CHECK19-NEXT: [[TMP17:%.*]] = sdiv exact i64 [[TMP16]], ptrtoint (ptr getelementptr (i8, ptr null, i32 1) to i64)2571// CHECK19-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[DOTOFFLOAD_SIZES]], ptr align 4 @.offload_sizes.1, i32 80, i1 false)2572// CHECK19-NEXT: [[TMP18:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 02573// CHECK19-NEXT: store ptr [[TMP6]], ptr [[TMP18]], align 42574// CHECK19-NEXT: [[TMP19:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 02575// CHECK19-NEXT: store ptr [[TMP6]], ptr [[TMP19]], align 42576// CHECK19-NEXT: [[TMP20:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 02577// CHECK19-NEXT: store ptr null, ptr [[TMP20]], align 42578// CHECK19-NEXT: [[TMP21:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 12579// CHECK19-NEXT: store i32 [[TMP0]], ptr [[TMP21]], align 42580// CHECK19-NEXT: [[TMP22:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 12581// CHECK19-NEXT: store i32 [[TMP0]], ptr [[TMP22]], align 42582// CHECK19-NEXT: [[TMP23:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 12583// CHECK19-NEXT: store ptr null, ptr [[TMP23]], align 42584// CHECK19-NEXT: [[TMP24:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 22585// CHECK19-NEXT: store ptr [[TMP7]], ptr [[TMP24]], align 42586// CHECK19-NEXT: [[TMP25:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 22587// CHECK19-NEXT: store ptr [[TMP7]], ptr [[TMP25]], align 42588// CHECK19-NEXT: [[TMP26:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 22589// CHECK19-NEXT: store ptr null, ptr [[TMP26]], align 42590// CHECK19-NEXT: [[TMP27:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 32591// CHECK19-NEXT: store i32 [[TMP1]], ptr [[TMP27]], align 42592// CHECK19-NEXT: [[TMP28:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 32593// CHECK19-NEXT: store i32 [[TMP1]], ptr [[TMP28]], align 42594// CHECK19-NEXT: [[TMP29:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 32595// CHECK19-NEXT: store ptr null, ptr [[TMP29]], align 42596// CHECK19-NEXT: [[TMP30:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 42597// CHECK19-NEXT: store i32 [[TMP2]], ptr [[TMP30]], align 42598// CHECK19-NEXT: [[TMP31:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 42599// CHECK19-NEXT: store i32 [[TMP2]], ptr [[TMP31]], align 42600// CHECK19-NEXT: [[TMP32:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 42601// CHECK19-NEXT: store ptr null, ptr [[TMP32]], align 42602// CHECK19-NEXT: [[TMP33:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 52603// CHECK19-NEXT: store ptr [[VLA]], ptr [[TMP33]], align 42604// CHECK19-NEXT: [[TMP34:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 52605// CHECK19-NEXT: store ptr [[VLA]], ptr [[TMP34]], align 42606// CHECK19-NEXT: [[TMP35:%.*]] = getelementptr inbounds [10 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 52607// CHECK19-NEXT: store i64 [[TMP12]], ptr [[TMP35]], align 42608// CHECK19-NEXT: [[TMP36:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 52609// CHECK19-NEXT: store ptr null, ptr [[TMP36]], align 42610// CHECK19-NEXT: [[TMP37:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 62611// CHECK19-NEXT: store ptr [[THIS1]], ptr [[TMP37]], align 42612// CHECK19-NEXT: [[TMP38:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 62613// CHECK19-NEXT: store ptr [[A3]], ptr [[TMP38]], align 42614// CHECK19-NEXT: [[TMP39:%.*]] = getelementptr inbounds [10 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 62615// CHECK19-NEXT: store i64 [[TMP17]], ptr [[TMP39]], align 42616// CHECK19-NEXT: [[TMP40:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 62617// CHECK19-NEXT: store ptr null, ptr [[TMP40]], align 42618// CHECK19-NEXT: [[TMP41:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 72619// CHECK19-NEXT: store ptr [[THIS1]], ptr [[TMP41]], align 42620// CHECK19-NEXT: [[TMP42:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 72621// CHECK19-NEXT: store ptr [[B2]], ptr [[TMP42]], align 42622// CHECK19-NEXT: [[TMP43:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 72623// CHECK19-NEXT: store ptr null, ptr [[TMP43]], align 42624// CHECK19-NEXT: [[TMP44:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 82625// CHECK19-NEXT: store ptr [[THIS1]], ptr [[TMP44]], align 42626// CHECK19-NEXT: [[TMP45:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 82627// CHECK19-NEXT: store ptr [[A3]], ptr [[TMP45]], align 42628// CHECK19-NEXT: [[TMP46:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 82629// CHECK19-NEXT: store ptr null, ptr [[TMP46]], align 42630// CHECK19-NEXT: [[TMP47:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 92631// CHECK19-NEXT: store i32 [[TMP9]], ptr [[TMP47]], align 42632// CHECK19-NEXT: [[TMP48:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 92633// CHECK19-NEXT: store i32 [[TMP9]], ptr [[TMP48]], align 42634// CHECK19-NEXT: [[TMP49:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 92635// CHECK19-NEXT: store ptr null, ptr [[TMP49]], align 42636// CHECK19-NEXT: [[TMP50:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 02637// CHECK19-NEXT: [[TMP51:%.*]] = getelementptr inbounds [10 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 02638// CHECK19-NEXT: [[TMP52:%.*]] = getelementptr inbounds [10 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 02639// CHECK19-NEXT: [[TMP53:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 02640// CHECK19-NEXT: store i32 3, ptr [[TMP53]], align 42641// CHECK19-NEXT: [[TMP54:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12642// CHECK19-NEXT: store i32 10, ptr [[TMP54]], align 42643// CHECK19-NEXT: [[TMP55:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 22644// CHECK19-NEXT: store ptr [[TMP50]], ptr [[TMP55]], align 42645// CHECK19-NEXT: [[TMP56:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 32646// CHECK19-NEXT: store ptr [[TMP51]], ptr [[TMP56]], align 42647// CHECK19-NEXT: [[TMP57:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 42648// CHECK19-NEXT: store ptr [[TMP52]], ptr [[TMP57]], align 42649// CHECK19-NEXT: [[TMP58:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 52650// CHECK19-NEXT: store ptr @.offload_maptypes.2, ptr [[TMP58]], align 42651// CHECK19-NEXT: [[TMP59:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 62652// CHECK19-NEXT: store ptr null, ptr [[TMP59]], align 42653// CHECK19-NEXT: [[TMP60:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 72654// CHECK19-NEXT: store ptr null, ptr [[TMP60]], align 42655// CHECK19-NEXT: [[TMP61:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 82656// CHECK19-NEXT: store i64 0, ptr [[TMP61]], align 82657// CHECK19-NEXT: [[TMP62:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 92658// CHECK19-NEXT: store i64 0, ptr [[TMP62]], align 82659// CHECK19-NEXT: [[TMP63:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 102660// CHECK19-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP63]], align 42661// CHECK19-NEXT: [[TMP64:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 112662// CHECK19-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP64]], align 42663// CHECK19-NEXT: [[TMP65:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 122664// CHECK19-NEXT: store i32 0, ptr [[TMP65]], align 42665// CHECK19-NEXT: [[TMP66:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPe_l144.region_id, ptr [[KERNEL_ARGS]])2666// CHECK19-NEXT: [[TMP67:%.*]] = icmp ne i32 [[TMP66]], 02667// CHECK19-NEXT: br i1 [[TMP67]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]2668// CHECK19: omp_offload.failed:2669// CHECK19-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPe_l144(ptr [[TMP6]], i32 [[TMP0]], ptr [[TMP7]], i32 [[TMP1]], i32 [[TMP2]], ptr [[VLA]], ptr [[THIS1]], i32 [[TMP9]]) #[[ATTR4]]2670// CHECK19-NEXT: br label [[OMP_OFFLOAD_CONT]]2671// CHECK19: omp_offload.cont:2672// CHECK19-NEXT: [[TMP68:%.*]] = load ptr, ptr [[SAVED_STACK]], align 42673// CHECK19-NEXT: call void @llvm.stackrestore.p0(ptr [[TMP68]])2674// CHECK19-NEXT: ret void2675//2676//2677// CHECK19-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPe_l1442678// CHECK19-SAME: (ptr [[S:%.*]], i32 [[VLA:%.*]], ptr [[VLA1:%.*]], i32 [[VLA2:%.*]], i32 [[VLA4:%.*]], ptr nonnull align 4 dereferenceable(8) [[VLA26:%.*]], ptr [[THIS:%.*]], i32 [[N:%.*]]) #[[ATTR2]] {2679// CHECK19-NEXT: entry:2680// CHECK19-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 42681// CHECK19-NEXT: [[VLA_ADDR:%.*]] = alloca i32, align 42682// CHECK19-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 42683// CHECK19-NEXT: [[VLA_ADDR3:%.*]] = alloca i32, align 42684// CHECK19-NEXT: [[VLA_ADDR5:%.*]] = alloca i32, align 42685// CHECK19-NEXT: [[VLA2_ADDR:%.*]] = alloca ptr, align 42686// CHECK19-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 42687// CHECK19-NEXT: [[N_ADDR:%.*]] = alloca i32, align 42688// CHECK19-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 42689// CHECK19-NEXT: store i32 [[VLA]], ptr [[VLA_ADDR]], align 42690// CHECK19-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 42691// CHECK19-NEXT: store i32 [[VLA2]], ptr [[VLA_ADDR3]], align 42692// CHECK19-NEXT: store i32 [[VLA4]], ptr [[VLA_ADDR5]], align 42693// CHECK19-NEXT: store ptr [[VLA26]], ptr [[VLA2_ADDR]], align 42694// CHECK19-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 42695// CHECK19-NEXT: store i32 [[N]], ptr [[N_ADDR]], align 42696// CHECK19-NEXT: [[TMP0:%.*]] = load i32, ptr [[VLA_ADDR]], align 42697// CHECK19-NEXT: [[TMP1:%.*]] = load i32, ptr [[VLA_ADDR3]], align 42698// CHECK19-NEXT: [[TMP2:%.*]] = load i32, ptr [[VLA_ADDR5]], align 42699// CHECK19-NEXT: [[TMP3:%.*]] = load ptr, ptr [[VLA2_ADDR]], align 4, !nonnull [[META8]], !align [[META9]]2700// CHECK19-NEXT: [[TMP4:%.*]] = load ptr, ptr [[THIS_ADDR]], align 42701// CHECK19-NEXT: [[TMP5:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 42702// CHECK19-NEXT: [[TMP6:%.*]] = load ptr, ptr [[S_ADDR]], align 42703// CHECK19-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB1]], i32 8, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPe_l144.omp_outlined, i32 [[TMP0]], ptr [[TMP5]], ptr [[TMP4]], i32 [[TMP1]], i32 [[TMP2]], ptr [[TMP3]], ptr [[N_ADDR]], ptr [[TMP6]])2704// CHECK19-NEXT: ret void2705//2706//2707// CHECK19-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2St7St_funcEPS_iPe_l144.omp_outlined2708// CHECK19-SAME: (ptr noalias [[DOTGLOBAL_TID_:%.*]], ptr noalias [[DOTBOUND_TID_:%.*]], i32 [[VLA:%.*]], ptr [[VLA1:%.*]], ptr [[THIS:%.*]], i32 [[VLA2:%.*]], i32 [[VLA4:%.*]], ptr nonnull align 4 dereferenceable(8) [[VLA26:%.*]], ptr nonnull align 4 dereferenceable(4) [[N:%.*]], ptr [[S:%.*]]) #[[ATTR2]] {2709// CHECK19-NEXT: entry:2710// CHECK19-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 42711// CHECK19-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 42712// CHECK19-NEXT: [[VLA_ADDR:%.*]] = alloca i32, align 42713// CHECK19-NEXT: [[VLA1_ADDR:%.*]] = alloca ptr, align 42714// CHECK19-NEXT: [[THIS_ADDR:%.*]] = alloca ptr, align 42715// CHECK19-NEXT: [[VLA_ADDR3:%.*]] = alloca i32, align 42716// CHECK19-NEXT: [[VLA_ADDR5:%.*]] = alloca i32, align 42717// CHECK19-NEXT: [[VLA2_ADDR:%.*]] = alloca ptr, align 42718// CHECK19-NEXT: [[N_ADDR:%.*]] = alloca ptr, align 42719// CHECK19-NEXT: [[S_ADDR:%.*]] = alloca ptr, align 42720// CHECK19-NEXT: [[SAVED_STACK:%.*]] = alloca ptr, align 42721// CHECK19-NEXT: [[__VLA_EXPR0:%.*]] = alloca i32, align 42722// CHECK19-NEXT: [[__VLA_EXPR1:%.*]] = alloca i32, align 42723// CHECK19-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 42724// CHECK19-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 42725// CHECK19-NEXT: store i32 [[VLA]], ptr [[VLA_ADDR]], align 42726// CHECK19-NEXT: store ptr [[VLA1]], ptr [[VLA1_ADDR]], align 42727// CHECK19-NEXT: store ptr [[THIS]], ptr [[THIS_ADDR]], align 42728// CHECK19-NEXT: store i32 [[VLA2]], ptr [[VLA_ADDR3]], align 42729// CHECK19-NEXT: store i32 [[VLA4]], ptr [[VLA_ADDR5]], align 42730// CHECK19-NEXT: store ptr [[VLA26]], ptr [[VLA2_ADDR]], align 42731// CHECK19-NEXT: store ptr [[N]], ptr [[N_ADDR]], align 42732// CHECK19-NEXT: store ptr [[S]], ptr [[S_ADDR]], align 42733// CHECK19-NEXT: [[TMP0:%.*]] = load i32, ptr [[VLA_ADDR]], align 42734// CHECK19-NEXT: [[TMP1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 42735// CHECK19-NEXT: [[TMP2:%.*]] = load i32, ptr [[VLA_ADDR3]], align 42736// CHECK19-NEXT: [[TMP3:%.*]] = load i32, ptr [[VLA_ADDR5]], align 42737// CHECK19-NEXT: [[TMP4:%.*]] = load ptr, ptr [[VLA2_ADDR]], align 4, !nonnull [[META8]], !align [[META9]]2738// CHECK19-NEXT: [[TMP5:%.*]] = load ptr, ptr [[N_ADDR]], align 4, !nonnull [[META8]], !align [[META9]]2739// CHECK19-NEXT: [[TMP6:%.*]] = call ptr @llvm.stacksave.p0()2740// CHECK19-NEXT: store ptr [[TMP6]], ptr [[SAVED_STACK]], align 42741// CHECK19-NEXT: [[TMP7:%.*]] = mul nuw i32 [[TMP2]], [[TMP3]]2742// CHECK19-NEXT: [[VLA7:%.*]] = alloca double, i32 [[TMP7]], align 1282743// CHECK19-NEXT: store i32 [[TMP2]], ptr [[__VLA_EXPR0]], align 42744// CHECK19-NEXT: store i32 [[TMP3]], ptr [[__VLA_EXPR1]], align 42745// CHECK19-NEXT: [[TMP8:%.*]] = mul nuw i32 [[TMP2]], [[TMP3]]2746// CHECK19-NEXT: [[TMP9:%.*]] = mul nuw i32 [[TMP8]], 82747// CHECK19-NEXT: call void @llvm.memcpy.p0.p0.i32(ptr align 128 [[VLA7]], ptr align 128 [[TMP4]], i32 [[TMP9]], i1 false)2748// CHECK19-NEXT: [[B:%.*]] = getelementptr inbounds nuw [[STRUCT_ST:%.*]], ptr [[TMP1]], i32 0, i32 12749// CHECK19-NEXT: [[TMP10:%.*]] = load i32, ptr [[B]], align 42750// CHECK19-NEXT: [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[TMP1]], i32 0, i32 02751// CHECK19-NEXT: store i32 [[TMP10]], ptr [[A]], align 42752// CHECK19-NEXT: [[CONV:%.*]] = sitofp i32 [[TMP10]] to double2753// CHECK19-NEXT: [[TMP11:%.*]] = mul nsw i32 1, [[TMP3]]2754// CHECK19-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds double, ptr [[VLA7]], i32 [[TMP11]]2755// CHECK19-NEXT: [[TMP12:%.*]] = load i32, ptr [[TMP5]], align 42756// CHECK19-NEXT: [[SUB:%.*]] = sub nsw i32 [[TMP12]], 12757// CHECK19-NEXT: [[ARRAYIDX8:%.*]] = getelementptr inbounds double, ptr [[ARRAYIDX]], i32 [[SUB]]2758// CHECK19-NEXT: store double [[CONV]], ptr [[ARRAYIDX8]], align 82759// CHECK19-NEXT: [[CONV9:%.*]] = fpext double [[CONV]] to x86_fp802760// CHECK19-NEXT: [[TMP13:%.*]] = load ptr, ptr [[VLA1_ADDR]], align 42761// CHECK19-NEXT: [[B10:%.*]] = getelementptr inbounds nuw [[STRUCT_ST]], ptr [[TMP1]], i32 0, i32 12762// CHECK19-NEXT: [[TMP14:%.*]] = load i32, ptr [[B10]], align 42763// CHECK19-NEXT: [[ARRAYIDX11:%.*]] = getelementptr inbounds x86_fp80, ptr [[TMP13]], i32 [[TMP14]]2764// CHECK19-NEXT: store x86_fp80 [[CONV9]], ptr [[ARRAYIDX11]], align 42765// CHECK19-NEXT: [[TMP15:%.*]] = load ptr, ptr [[SAVED_STACK]], align 42766// CHECK19-NEXT: call void @llvm.stackrestore.p0(ptr [[TMP15]])2767// CHECK19-NEXT: ret void2768//2769