1180 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// RUN: %clang_cc1 -DCHECK -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix=CHECK13// RUN: %clang_cc1 -DCHECK -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s4// RUN: %clang_cc1 -DCHECK -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix=CHECK15// RUN: %clang_cc1 -DCHECK -verify -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix=CHECK36// RUN: %clang_cc1 -DCHECK -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s7// RUN: %clang_cc1 -DCHECK -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix=CHECK38 9// RUN: %clang_cc1 -DCHECK -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"10// RUN: %clang_cc1 -DCHECK -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s11// RUN: %clang_cc1 -DCHECK -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"12// RUN: %clang_cc1 -DCHECK -verify -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"13// RUN: %clang_cc1 -DCHECK -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s14// RUN: %clang_cc1 -DCHECK -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"15 16// RUN: %clang_cc1 -DLAMBDA -verify -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix=CHECK917// RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s18// RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix=CHECK919 20// RUN: %clang_cc1 -DLAMBDA -verify -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"21// RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s22// RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"23 24// expected-no-diagnostics25#ifndef HEADER26#define HEADER27 28int x;29#pragma omp threadprivate(x)30 31template <typename T>32T tmain() {33 int a[2];34#pragma omp target35#pragma omp teams distribute parallel for copyin(x)36 for (int i = 0; i < 2; ++i) {37 a[i] = x;38 }39 return T();40}41 42int main() {43 int a[2];44#ifdef LAMBDA45 [&]() {46#pragma omp target47#pragma omp teams distribute parallel for copyin(x)48 for (int i = 0; i < 2; ++i) {49 50 // Skip global, bound tid and loop vars51 a[i] = x;52 53 54 // Skip global, bound tid and loop vars55 56 57 [&]() {58 a[i] = x;59 }();60 }61 }();62 return 0;63#else64#pragma omp target65#pragma omp teams distribute parallel for copyin(x)66 for (int i = 0; i < 2; ++i) {67 a[i] = x;68 }69 return tmain<int>();70 //return 0;71#endif72}73 74 75 76// Skip global, bound tid and loop vars77 78 79// Skip global, bound tid and loop vars80 81 82 83 84// Skip global, bound tid and loop vars85 86 87// Skip global, bound tid and loop vars88// prev lb and ub89 90// iter variables91 92 93 94#endif95// CHECK1-LABEL: define {{[^@]+}}@main96// CHECK1-SAME: () #[[ATTR0:[0-9]+]] {97// CHECK1-NEXT: entry:98// CHECK1-NEXT: [[RETVAL:%.*]] = alloca i32, align 499// CHECK1-NEXT: [[A:%.*]] = alloca [2 x i32], align 4100// CHECK1-NEXT: [[X_CASTED:%.*]] = alloca i64, align 8101// CHECK1-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [2 x ptr], align 8102// CHECK1-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [2 x ptr], align 8103// CHECK1-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [2 x ptr], align 8104// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4105// CHECK1-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8106// CHECK1-NEXT: store i32 0, ptr [[RETVAL]], align 4107// CHECK1-NEXT: [[TMP0:%.*]] = call align 4 ptr @llvm.threadlocal.address.p0(ptr align 4 @x)108// CHECK1-NEXT: [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4109// CHECK1-NEXT: store i32 [[TMP1]], ptr [[X_CASTED]], align 4110// CHECK1-NEXT: [[TMP2:%.*]] = load i64, ptr [[X_CASTED]], align 8111// CHECK1-NEXT: [[TMP3:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0112// CHECK1-NEXT: store i64 [[TMP2]], ptr [[TMP3]], align 8113// CHECK1-NEXT: [[TMP4:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0114// CHECK1-NEXT: store i64 [[TMP2]], ptr [[TMP4]], align 8115// CHECK1-NEXT: [[TMP5:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0116// CHECK1-NEXT: store ptr null, ptr [[TMP5]], align 8117// CHECK1-NEXT: [[TMP6:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1118// CHECK1-NEXT: store ptr [[A]], ptr [[TMP6]], align 8119// CHECK1-NEXT: [[TMP7:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1120// CHECK1-NEXT: store ptr [[A]], ptr [[TMP7]], align 8121// CHECK1-NEXT: [[TMP8:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1122// CHECK1-NEXT: store ptr null, ptr [[TMP8]], align 8123// CHECK1-NEXT: [[TMP9:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0124// CHECK1-NEXT: [[TMP10:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0125// CHECK1-NEXT: [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0126// CHECK1-NEXT: store i32 3, ptr [[TMP11]], align 4127// CHECK1-NEXT: [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1128// CHECK1-NEXT: store i32 2, ptr [[TMP12]], align 4129// CHECK1-NEXT: [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2130// CHECK1-NEXT: store ptr [[TMP9]], ptr [[TMP13]], align 8131// CHECK1-NEXT: [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3132// CHECK1-NEXT: store ptr [[TMP10]], ptr [[TMP14]], align 8133// CHECK1-NEXT: [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4134// CHECK1-NEXT: store ptr @.offload_sizes, ptr [[TMP15]], align 8135// CHECK1-NEXT: [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5136// CHECK1-NEXT: store ptr @.offload_maptypes, ptr [[TMP16]], align 8137// CHECK1-NEXT: [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6138// CHECK1-NEXT: store ptr null, ptr [[TMP17]], align 8139// CHECK1-NEXT: [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7140// CHECK1-NEXT: store ptr null, ptr [[TMP18]], align 8141// CHECK1-NEXT: [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8142// CHECK1-NEXT: store i64 2, ptr [[TMP19]], align 8143// CHECK1-NEXT: [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9144// CHECK1-NEXT: store i64 0, ptr [[TMP20]], align 8145// CHECK1-NEXT: [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10146// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP21]], align 4147// CHECK1-NEXT: [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11148// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP22]], align 4149// CHECK1-NEXT: [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12150// CHECK1-NEXT: store i32 0, ptr [[TMP23]], align 4151// CHECK1-NEXT: [[TMP24:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB3:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.region_id, ptr [[KERNEL_ARGS]])152// CHECK1-NEXT: [[TMP25:%.*]] = icmp ne i32 [[TMP24]], 0153// CHECK1-NEXT: br i1 [[TMP25]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]154// CHECK1: omp_offload.failed:155// CHECK1-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64(i64 [[TMP2]], ptr [[A]]) #[[ATTR2:[0-9]+]]156// CHECK1-NEXT: br label [[OMP_OFFLOAD_CONT]]157// CHECK1: omp_offload.cont:158// CHECK1-NEXT: [[CALL:%.*]] = call noundef signext i32 @_Z5tmainIiET_v()159// CHECK1-NEXT: ret i32 [[CALL]]160//161//162// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64163// CHECK1-SAME: (i64 noundef [[X:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]]) #[[ATTR1:[0-9]+]] {164// CHECK1-NEXT: entry:165// CHECK1-NEXT: [[X_ADDR:%.*]] = alloca i64, align 8166// CHECK1-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 8167// CHECK1-NEXT: store i64 [[X]], ptr [[X_ADDR]], align 8168// CHECK1-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 8169// CHECK1-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 8170// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB3]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.omp_outlined, ptr [[TMP0]], ptr [[X_ADDR]])171// CHECK1-NEXT: ret void172//173//174// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.omp_outlined175// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR1]] {176// CHECK1-NEXT: entry:177// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8178// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8179// CHECK1-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 8180// CHECK1-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 8181// CHECK1-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4182// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4183// CHECK1-NEXT: [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4184// CHECK1-NEXT: [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4185// CHECK1-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4186// CHECK1-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4187// CHECK1-NEXT: [[I:%.*]] = alloca i32, align 4188// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8189// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8190// CHECK1-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 8191// CHECK1-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 8192// CHECK1-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 8193// CHECK1-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 8194// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_COMB_LB]], align 4195// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_COMB_UB]], align 4196// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4197// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4198// CHECK1-NEXT: [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8199// CHECK1-NEXT: [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 4200// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP3]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)201// CHECK1-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4202// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 1203// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]204// CHECK1: cond.true:205// CHECK1-NEXT: br label [[COND_END:%.*]]206// CHECK1: cond.false:207// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4208// CHECK1-NEXT: br label [[COND_END]]209// CHECK1: cond.end:210// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP5]], [[COND_FALSE]] ]211// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4212// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4213// CHECK1-NEXT: store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 4214// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]215// CHECK1: omp.inner.for.cond:216// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4217// CHECK1-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4218// CHECK1-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]219// CHECK1-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]220// CHECK1: omp.inner.for.body:221// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4222// CHECK1-NEXT: [[TMP10:%.*]] = zext i32 [[TMP9]] to i64223// CHECK1-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4224// CHECK1-NEXT: [[TMP12:%.*]] = zext i32 [[TMP11]] to i64225// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB3]], i32 4, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.omp_outlined.omp_outlined, i64 [[TMP10]], i64 [[TMP12]], ptr [[TMP0]], ptr [[TMP1]])226// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]227// CHECK1: omp.inner.for.inc:228// CHECK1-NEXT: [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4229// CHECK1-NEXT: [[TMP14:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4230// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP13]], [[TMP14]]231// CHECK1-NEXT: store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4232// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]233// CHECK1: omp.inner.for.end:234// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]235// CHECK1: omp.loop.exit:236// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP3]])237// CHECK1-NEXT: ret void238//239//240// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.omp_outlined.omp_outlined241// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], i64 noundef [[DOTPREVIOUS_LB_:%.*]], i64 noundef [[DOTPREVIOUS_UB_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR1]] {242// CHECK1-NEXT: entry:243// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8244// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8245// CHECK1-NEXT: [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 8246// CHECK1-NEXT: [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 8247// CHECK1-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 8248// CHECK1-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 8249// CHECK1-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4250// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4251// CHECK1-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 4252// CHECK1-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 4253// CHECK1-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4254// CHECK1-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4255// CHECK1-NEXT: [[I:%.*]] = alloca i32, align 4256// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8257// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8258// CHECK1-NEXT: store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 8259// CHECK1-NEXT: store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 8260// CHECK1-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 8261// CHECK1-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 8262// CHECK1-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 8263// CHECK1-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 8264// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_LB]], align 4265// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_UB]], align 4266// CHECK1-NEXT: [[TMP2:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 8267// CHECK1-NEXT: [[CONV:%.*]] = trunc i64 [[TMP2]] to i32268// CHECK1-NEXT: [[TMP3:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 8269// CHECK1-NEXT: [[CONV1:%.*]] = trunc i64 [[TMP3]] to i32270// CHECK1-NEXT: store i32 [[CONV]], ptr [[DOTOMP_LB]], align 4271// CHECK1-NEXT: store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 4272// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4273// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4274// CHECK1-NEXT: [[TMP4:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8275// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[TMP4]], align 4276// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB2:[0-9]+]], i32 [[TMP5]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)277// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4278// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP6]], 1279// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]280// CHECK1: cond.true:281// CHECK1-NEXT: br label [[COND_END:%.*]]282// CHECK1: cond.false:283// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4284// CHECK1-NEXT: br label [[COND_END]]285// CHECK1: cond.end:286// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP7]], [[COND_FALSE]] ]287// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_UB]], align 4288// CHECK1-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4289// CHECK1-NEXT: store i32 [[TMP8]], ptr [[DOTOMP_IV]], align 4290// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]291// CHECK1: omp.inner.for.cond:292// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4293// CHECK1-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4294// CHECK1-NEXT: [[CMP2:%.*]] = icmp sle i32 [[TMP9]], [[TMP10]]295// CHECK1-NEXT: br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]296// CHECK1: omp.inner.for.body:297// CHECK1-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4298// CHECK1-NEXT: [[MUL:%.*]] = mul nsw i32 [[TMP11]], 1299// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 0, [[MUL]]300// CHECK1-NEXT: store i32 [[ADD]], ptr [[I]], align 4301// CHECK1-NEXT: [[TMP12:%.*]] = load i32, ptr [[TMP1]], align 4302// CHECK1-NEXT: [[TMP13:%.*]] = load i32, ptr [[I]], align 4303// CHECK1-NEXT: [[IDXPROM:%.*]] = sext i32 [[TMP13]] to i64304// CHECK1-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[TMP0]], i64 0, i64 [[IDXPROM]]305// CHECK1-NEXT: store i32 [[TMP12]], ptr [[ARRAYIDX]], align 4306// CHECK1-NEXT: br label [[OMP_BODY_CONTINUE:%.*]]307// CHECK1: omp.body.continue:308// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]309// CHECK1: omp.inner.for.inc:310// CHECK1-NEXT: [[TMP14:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4311// CHECK1-NEXT: [[ADD3:%.*]] = add nsw i32 [[TMP14]], 1312// CHECK1-NEXT: store i32 [[ADD3]], ptr [[DOTOMP_IV]], align 4313// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]314// CHECK1: omp.inner.for.end:315// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]316// CHECK1: omp.loop.exit:317// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP5]])318// CHECK1-NEXT: ret void319//320//321// CHECK1-LABEL: define {{[^@]+}}@_Z5tmainIiET_v322// CHECK1-SAME: () #[[ATTR4:[0-9]+]] comdat {323// CHECK1-NEXT: entry:324// CHECK1-NEXT: [[A:%.*]] = alloca [2 x i32], align 4325// CHECK1-NEXT: [[X_CASTED:%.*]] = alloca i64, align 8326// CHECK1-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [2 x ptr], align 8327// CHECK1-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [2 x ptr], align 8328// CHECK1-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [2 x ptr], align 8329// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4330// CHECK1-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8331// CHECK1-NEXT: [[TMP0:%.*]] = call align 4 ptr @llvm.threadlocal.address.p0(ptr align 4 @x)332// CHECK1-NEXT: [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4333// CHECK1-NEXT: store i32 [[TMP1]], ptr [[X_CASTED]], align 4334// CHECK1-NEXT: [[TMP2:%.*]] = load i64, ptr [[X_CASTED]], align 8335// CHECK1-NEXT: [[TMP3:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0336// CHECK1-NEXT: store i64 [[TMP2]], ptr [[TMP3]], align 8337// CHECK1-NEXT: [[TMP4:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0338// CHECK1-NEXT: store i64 [[TMP2]], ptr [[TMP4]], align 8339// CHECK1-NEXT: [[TMP5:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0340// CHECK1-NEXT: store ptr null, ptr [[TMP5]], align 8341// CHECK1-NEXT: [[TMP6:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1342// CHECK1-NEXT: store ptr [[A]], ptr [[TMP6]], align 8343// CHECK1-NEXT: [[TMP7:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1344// CHECK1-NEXT: store ptr [[A]], ptr [[TMP7]], align 8345// CHECK1-NEXT: [[TMP8:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1346// CHECK1-NEXT: store ptr null, ptr [[TMP8]], align 8347// CHECK1-NEXT: [[TMP9:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0348// CHECK1-NEXT: [[TMP10:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0349// CHECK1-NEXT: [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0350// CHECK1-NEXT: store i32 3, ptr [[TMP11]], align 4351// CHECK1-NEXT: [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1352// CHECK1-NEXT: store i32 2, ptr [[TMP12]], align 4353// CHECK1-NEXT: [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2354// CHECK1-NEXT: store ptr [[TMP9]], ptr [[TMP13]], align 8355// CHECK1-NEXT: [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3356// CHECK1-NEXT: store ptr [[TMP10]], ptr [[TMP14]], align 8357// CHECK1-NEXT: [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4358// CHECK1-NEXT: store ptr @.offload_sizes.1, ptr [[TMP15]], align 8359// CHECK1-NEXT: [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5360// CHECK1-NEXT: store ptr @.offload_maptypes.2, ptr [[TMP16]], align 8361// CHECK1-NEXT: [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6362// CHECK1-NEXT: store ptr null, ptr [[TMP17]], align 8363// CHECK1-NEXT: [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7364// CHECK1-NEXT: store ptr null, ptr [[TMP18]], align 8365// CHECK1-NEXT: [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8366// CHECK1-NEXT: store i64 2, ptr [[TMP19]], align 8367// CHECK1-NEXT: [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9368// CHECK1-NEXT: store i64 0, ptr [[TMP20]], align 8369// CHECK1-NEXT: [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10370// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP21]], align 4371// CHECK1-NEXT: [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11372// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP22]], align 4373// CHECK1-NEXT: [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12374// CHECK1-NEXT: store i32 0, ptr [[TMP23]], align 4375// CHECK1-NEXT: [[TMP24:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB3]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.region_id, ptr [[KERNEL_ARGS]])376// CHECK1-NEXT: [[TMP25:%.*]] = icmp ne i32 [[TMP24]], 0377// CHECK1-NEXT: br i1 [[TMP25]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]378// CHECK1: omp_offload.failed:379// CHECK1-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34(i64 [[TMP2]], ptr [[A]]) #[[ATTR2]]380// CHECK1-NEXT: br label [[OMP_OFFLOAD_CONT]]381// CHECK1: omp_offload.cont:382// CHECK1-NEXT: ret i32 0383//384//385// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34386// CHECK1-SAME: (i64 noundef [[X:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]]) #[[ATTR1]] {387// CHECK1-NEXT: entry:388// CHECK1-NEXT: [[X_ADDR:%.*]] = alloca i64, align 8389// CHECK1-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 8390// CHECK1-NEXT: store i64 [[X]], ptr [[X_ADDR]], align 8391// CHECK1-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 8392// CHECK1-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 8393// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB3]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.omp_outlined, ptr [[TMP0]], ptr [[X_ADDR]])394// CHECK1-NEXT: ret void395//396//397// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.omp_outlined398// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR1]] {399// CHECK1-NEXT: entry:400// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8401// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8402// CHECK1-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 8403// CHECK1-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 8404// CHECK1-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4405// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4406// CHECK1-NEXT: [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4407// CHECK1-NEXT: [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4408// CHECK1-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4409// CHECK1-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4410// CHECK1-NEXT: [[I:%.*]] = alloca i32, align 4411// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8412// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8413// CHECK1-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 8414// CHECK1-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 8415// CHECK1-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 8416// CHECK1-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 8417// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_COMB_LB]], align 4418// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_COMB_UB]], align 4419// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4420// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4421// CHECK1-NEXT: [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8422// CHECK1-NEXT: [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 4423// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP3]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)424// CHECK1-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4425// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 1426// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]427// CHECK1: cond.true:428// CHECK1-NEXT: br label [[COND_END:%.*]]429// CHECK1: cond.false:430// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4431// CHECK1-NEXT: br label [[COND_END]]432// CHECK1: cond.end:433// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP5]], [[COND_FALSE]] ]434// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4435// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4436// CHECK1-NEXT: store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 4437// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]438// CHECK1: omp.inner.for.cond:439// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4440// CHECK1-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4441// CHECK1-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]442// CHECK1-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]443// CHECK1: omp.inner.for.body:444// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4445// CHECK1-NEXT: [[TMP10:%.*]] = zext i32 [[TMP9]] to i64446// CHECK1-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4447// CHECK1-NEXT: [[TMP12:%.*]] = zext i32 [[TMP11]] to i64448// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB3]], i32 4, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.omp_outlined.omp_outlined, i64 [[TMP10]], i64 [[TMP12]], ptr [[TMP0]], ptr [[TMP1]])449// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]450// CHECK1: omp.inner.for.inc:451// CHECK1-NEXT: [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4452// CHECK1-NEXT: [[TMP14:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4453// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP13]], [[TMP14]]454// CHECK1-NEXT: store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4455// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]456// CHECK1: omp.inner.for.end:457// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]458// CHECK1: omp.loop.exit:459// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP3]])460// CHECK1-NEXT: ret void461//462//463// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.omp_outlined.omp_outlined464// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], i64 noundef [[DOTPREVIOUS_LB_:%.*]], i64 noundef [[DOTPREVIOUS_UB_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR1]] {465// CHECK1-NEXT: entry:466// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8467// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8468// CHECK1-NEXT: [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 8469// CHECK1-NEXT: [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 8470// CHECK1-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 8471// CHECK1-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 8472// CHECK1-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4473// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4474// CHECK1-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 4475// CHECK1-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 4476// CHECK1-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4477// CHECK1-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4478// CHECK1-NEXT: [[I:%.*]] = alloca i32, align 4479// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8480// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8481// CHECK1-NEXT: store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 8482// CHECK1-NEXT: store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 8483// CHECK1-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 8484// CHECK1-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 8485// CHECK1-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 8486// CHECK1-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 8487// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_LB]], align 4488// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_UB]], align 4489// CHECK1-NEXT: [[TMP2:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 8490// CHECK1-NEXT: [[CONV:%.*]] = trunc i64 [[TMP2]] to i32491// CHECK1-NEXT: [[TMP3:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 8492// CHECK1-NEXT: [[CONV1:%.*]] = trunc i64 [[TMP3]] to i32493// CHECK1-NEXT: store i32 [[CONV]], ptr [[DOTOMP_LB]], align 4494// CHECK1-NEXT: store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 4495// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4496// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4497// CHECK1-NEXT: [[TMP4:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8498// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[TMP4]], align 4499// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB2]], i32 [[TMP5]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)500// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4501// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP6]], 1502// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]503// CHECK1: cond.true:504// CHECK1-NEXT: br label [[COND_END:%.*]]505// CHECK1: cond.false:506// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4507// CHECK1-NEXT: br label [[COND_END]]508// CHECK1: cond.end:509// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP7]], [[COND_FALSE]] ]510// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_UB]], align 4511// CHECK1-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4512// CHECK1-NEXT: store i32 [[TMP8]], ptr [[DOTOMP_IV]], align 4513// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]514// CHECK1: omp.inner.for.cond:515// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4516// CHECK1-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4517// CHECK1-NEXT: [[CMP2:%.*]] = icmp sle i32 [[TMP9]], [[TMP10]]518// CHECK1-NEXT: br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]519// CHECK1: omp.inner.for.body:520// CHECK1-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4521// CHECK1-NEXT: [[MUL:%.*]] = mul nsw i32 [[TMP11]], 1522// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 0, [[MUL]]523// CHECK1-NEXT: store i32 [[ADD]], ptr [[I]], align 4524// CHECK1-NEXT: [[TMP12:%.*]] = load i32, ptr [[TMP1]], align 4525// CHECK1-NEXT: [[TMP13:%.*]] = load i32, ptr [[I]], align 4526// CHECK1-NEXT: [[IDXPROM:%.*]] = sext i32 [[TMP13]] to i64527// CHECK1-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[TMP0]], i64 0, i64 [[IDXPROM]]528// CHECK1-NEXT: store i32 [[TMP12]], ptr [[ARRAYIDX]], align 4529// CHECK1-NEXT: br label [[OMP_BODY_CONTINUE:%.*]]530// CHECK1: omp.body.continue:531// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]532// CHECK1: omp.inner.for.inc:533// CHECK1-NEXT: [[TMP14:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4534// CHECK1-NEXT: [[ADD3:%.*]] = add nsw i32 [[TMP14]], 1535// CHECK1-NEXT: store i32 [[ADD3]], ptr [[DOTOMP_IV]], align 4536// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]537// CHECK1: omp.inner.for.end:538// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]539// CHECK1: omp.loop.exit:540// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP5]])541// CHECK1-NEXT: ret void542//543//544// CHECK1-LABEL: define {{[^@]+}}@_ZTW1x545// CHECK1-SAME: () #[[ATTR5:[0-9]+]] comdat {546// CHECK1-NEXT: [[TMP1:%.*]] = call align 4 ptr @llvm.threadlocal.address.p0(ptr align 4 @x)547// CHECK1-NEXT: ret ptr [[TMP1]]548//549//550// CHECK3-LABEL: define {{[^@]+}}@main551// CHECK3-SAME: () #[[ATTR0:[0-9]+]] {552// CHECK3-NEXT: entry:553// CHECK3-NEXT: [[RETVAL:%.*]] = alloca i32, align 4554// CHECK3-NEXT: [[A:%.*]] = alloca [2 x i32], align 4555// CHECK3-NEXT: [[X_CASTED:%.*]] = alloca i32, align 4556// CHECK3-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [2 x ptr], align 4557// CHECK3-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [2 x ptr], align 4558// CHECK3-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [2 x ptr], align 4559// CHECK3-NEXT: [[TMP:%.*]] = alloca i32, align 4560// CHECK3-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8561// CHECK3-NEXT: store i32 0, ptr [[RETVAL]], align 4562// CHECK3-NEXT: [[TMP0:%.*]] = call align 4 ptr @llvm.threadlocal.address.p0(ptr align 4 @x)563// CHECK3-NEXT: [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4564// CHECK3-NEXT: store i32 [[TMP1]], ptr [[X_CASTED]], align 4565// CHECK3-NEXT: [[TMP2:%.*]] = load i32, ptr [[X_CASTED]], align 4566// CHECK3-NEXT: [[TMP3:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0567// CHECK3-NEXT: store i32 [[TMP2]], ptr [[TMP3]], align 4568// CHECK3-NEXT: [[TMP4:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0569// CHECK3-NEXT: store i32 [[TMP2]], ptr [[TMP4]], align 4570// CHECK3-NEXT: [[TMP5:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0571// CHECK3-NEXT: store ptr null, ptr [[TMP5]], align 4572// CHECK3-NEXT: [[TMP6:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1573// CHECK3-NEXT: store ptr [[A]], ptr [[TMP6]], align 4574// CHECK3-NEXT: [[TMP7:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1575// CHECK3-NEXT: store ptr [[A]], ptr [[TMP7]], align 4576// CHECK3-NEXT: [[TMP8:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 1577// CHECK3-NEXT: store ptr null, ptr [[TMP8]], align 4578// CHECK3-NEXT: [[TMP9:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0579// CHECK3-NEXT: [[TMP10:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0580// CHECK3-NEXT: [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0581// CHECK3-NEXT: store i32 3, ptr [[TMP11]], align 4582// CHECK3-NEXT: [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1583// CHECK3-NEXT: store i32 2, ptr [[TMP12]], align 4584// CHECK3-NEXT: [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2585// CHECK3-NEXT: store ptr [[TMP9]], ptr [[TMP13]], align 4586// CHECK3-NEXT: [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3587// CHECK3-NEXT: store ptr [[TMP10]], ptr [[TMP14]], align 4588// CHECK3-NEXT: [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4589// CHECK3-NEXT: store ptr @.offload_sizes, ptr [[TMP15]], align 4590// CHECK3-NEXT: [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5591// CHECK3-NEXT: store ptr @.offload_maptypes, ptr [[TMP16]], align 4592// CHECK3-NEXT: [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6593// CHECK3-NEXT: store ptr null, ptr [[TMP17]], align 4594// CHECK3-NEXT: [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7595// CHECK3-NEXT: store ptr null, ptr [[TMP18]], align 4596// CHECK3-NEXT: [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8597// CHECK3-NEXT: store i64 2, ptr [[TMP19]], align 8598// CHECK3-NEXT: [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9599// CHECK3-NEXT: store i64 0, ptr [[TMP20]], align 8600// CHECK3-NEXT: [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10601// CHECK3-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP21]], align 4602// CHECK3-NEXT: [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11603// CHECK3-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP22]], align 4604// CHECK3-NEXT: [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12605// CHECK3-NEXT: store i32 0, ptr [[TMP23]], align 4606// CHECK3-NEXT: [[TMP24:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB3:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.region_id, ptr [[KERNEL_ARGS]])607// CHECK3-NEXT: [[TMP25:%.*]] = icmp ne i32 [[TMP24]], 0608// CHECK3-NEXT: br i1 [[TMP25]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]609// CHECK3: omp_offload.failed:610// CHECK3-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64(i32 [[TMP2]], ptr [[A]]) #[[ATTR2:[0-9]+]]611// CHECK3-NEXT: br label [[OMP_OFFLOAD_CONT]]612// CHECK3: omp_offload.cont:613// CHECK3-NEXT: [[CALL:%.*]] = call noundef i32 @_Z5tmainIiET_v()614// CHECK3-NEXT: ret i32 [[CALL]]615//616//617// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64618// CHECK3-SAME: (i32 noundef [[X:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]]) #[[ATTR1:[0-9]+]] {619// CHECK3-NEXT: entry:620// CHECK3-NEXT: [[X_ADDR:%.*]] = alloca i32, align 4621// CHECK3-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 4622// CHECK3-NEXT: store i32 [[X]], ptr [[X_ADDR]], align 4623// CHECK3-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 4624// CHECK3-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 4625// CHECK3-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB3]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.omp_outlined, ptr [[TMP0]], ptr [[X_ADDR]])626// CHECK3-NEXT: ret void627//628//629// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.omp_outlined630// CHECK3-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR1]] {631// CHECK3-NEXT: entry:632// CHECK3-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4633// CHECK3-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4634// CHECK3-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 4635// CHECK3-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 4636// CHECK3-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4637// CHECK3-NEXT: [[TMP:%.*]] = alloca i32, align 4638// CHECK3-NEXT: [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4639// CHECK3-NEXT: [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4640// CHECK3-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4641// CHECK3-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4642// CHECK3-NEXT: [[I:%.*]] = alloca i32, align 4643// CHECK3-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4644// CHECK3-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4645// CHECK3-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 4646// CHECK3-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 4647// CHECK3-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 4648// CHECK3-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 4649// CHECK3-NEXT: store i32 0, ptr [[DOTOMP_COMB_LB]], align 4650// CHECK3-NEXT: store i32 1, ptr [[DOTOMP_COMB_UB]], align 4651// CHECK3-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4652// CHECK3-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4653// CHECK3-NEXT: [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4654// CHECK3-NEXT: [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 4655// CHECK3-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP3]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)656// CHECK3-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4657// CHECK3-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 1658// CHECK3-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]659// CHECK3: cond.true:660// CHECK3-NEXT: br label [[COND_END:%.*]]661// CHECK3: cond.false:662// CHECK3-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4663// CHECK3-NEXT: br label [[COND_END]]664// CHECK3: cond.end:665// CHECK3-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP5]], [[COND_FALSE]] ]666// CHECK3-NEXT: store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4667// CHECK3-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4668// CHECK3-NEXT: store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 4669// CHECK3-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]670// CHECK3: omp.inner.for.cond:671// CHECK3-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4672// CHECK3-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4673// CHECK3-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]674// CHECK3-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]675// CHECK3: omp.inner.for.body:676// CHECK3-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4677// CHECK3-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4678// CHECK3-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB3]], i32 4, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.omp_outlined.omp_outlined, i32 [[TMP9]], i32 [[TMP10]], ptr [[TMP0]], ptr [[TMP1]])679// CHECK3-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]680// CHECK3: omp.inner.for.inc:681// CHECK3-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4682// CHECK3-NEXT: [[TMP12:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4683// CHECK3-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP11]], [[TMP12]]684// CHECK3-NEXT: store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4685// CHECK3-NEXT: br label [[OMP_INNER_FOR_COND]]686// CHECK3: omp.inner.for.end:687// CHECK3-NEXT: br label [[OMP_LOOP_EXIT:%.*]]688// CHECK3: omp.loop.exit:689// CHECK3-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP3]])690// CHECK3-NEXT: ret void691//692//693// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l64.omp_outlined.omp_outlined694// CHECK3-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], i32 noundef [[DOTPREVIOUS_LB_:%.*]], i32 noundef [[DOTPREVIOUS_UB_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR1]] {695// CHECK3-NEXT: entry:696// CHECK3-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4697// CHECK3-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4698// CHECK3-NEXT: [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i32, align 4699// CHECK3-NEXT: [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i32, align 4700// CHECK3-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 4701// CHECK3-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 4702// CHECK3-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4703// CHECK3-NEXT: [[TMP:%.*]] = alloca i32, align 4704// CHECK3-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 4705// CHECK3-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 4706// CHECK3-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4707// CHECK3-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4708// CHECK3-NEXT: [[I:%.*]] = alloca i32, align 4709// CHECK3-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4710// CHECK3-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4711// CHECK3-NEXT: store i32 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 4712// CHECK3-NEXT: store i32 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 4713// CHECK3-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 4714// CHECK3-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 4715// CHECK3-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 4716// CHECK3-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 4717// CHECK3-NEXT: store i32 0, ptr [[DOTOMP_LB]], align 4718// CHECK3-NEXT: store i32 1, ptr [[DOTOMP_UB]], align 4719// CHECK3-NEXT: [[TMP2:%.*]] = load i32, ptr [[DOTPREVIOUS_LB__ADDR]], align 4720// CHECK3-NEXT: [[TMP3:%.*]] = load i32, ptr [[DOTPREVIOUS_UB__ADDR]], align 4721// CHECK3-NEXT: store i32 [[TMP2]], ptr [[DOTOMP_LB]], align 4722// CHECK3-NEXT: store i32 [[TMP3]], ptr [[DOTOMP_UB]], align 4723// CHECK3-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4724// CHECK3-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4725// CHECK3-NEXT: [[TMP4:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4726// CHECK3-NEXT: [[TMP5:%.*]] = load i32, ptr [[TMP4]], align 4727// CHECK3-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB2:[0-9]+]], i32 [[TMP5]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)728// CHECK3-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4729// CHECK3-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP6]], 1730// CHECK3-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]731// CHECK3: cond.true:732// CHECK3-NEXT: br label [[COND_END:%.*]]733// CHECK3: cond.false:734// CHECK3-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4735// CHECK3-NEXT: br label [[COND_END]]736// CHECK3: cond.end:737// CHECK3-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP7]], [[COND_FALSE]] ]738// CHECK3-NEXT: store i32 [[COND]], ptr [[DOTOMP_UB]], align 4739// CHECK3-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4740// CHECK3-NEXT: store i32 [[TMP8]], ptr [[DOTOMP_IV]], align 4741// CHECK3-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]742// CHECK3: omp.inner.for.cond:743// CHECK3-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4744// CHECK3-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4745// CHECK3-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP9]], [[TMP10]]746// CHECK3-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]747// CHECK3: omp.inner.for.body:748// CHECK3-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4749// CHECK3-NEXT: [[MUL:%.*]] = mul nsw i32 [[TMP11]], 1750// CHECK3-NEXT: [[ADD:%.*]] = add nsw i32 0, [[MUL]]751// CHECK3-NEXT: store i32 [[ADD]], ptr [[I]], align 4752// CHECK3-NEXT: [[TMP12:%.*]] = load i32, ptr [[TMP1]], align 4753// CHECK3-NEXT: [[TMP13:%.*]] = load i32, ptr [[I]], align 4754// CHECK3-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[TMP0]], i32 0, i32 [[TMP13]]755// CHECK3-NEXT: store i32 [[TMP12]], ptr [[ARRAYIDX]], align 4756// CHECK3-NEXT: br label [[OMP_BODY_CONTINUE:%.*]]757// CHECK3: omp.body.continue:758// CHECK3-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]759// CHECK3: omp.inner.for.inc:760// CHECK3-NEXT: [[TMP14:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4761// CHECK3-NEXT: [[ADD2:%.*]] = add nsw i32 [[TMP14]], 1762// CHECK3-NEXT: store i32 [[ADD2]], ptr [[DOTOMP_IV]], align 4763// CHECK3-NEXT: br label [[OMP_INNER_FOR_COND]]764// CHECK3: omp.inner.for.end:765// CHECK3-NEXT: br label [[OMP_LOOP_EXIT:%.*]]766// CHECK3: omp.loop.exit:767// CHECK3-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP5]])768// CHECK3-NEXT: ret void769//770//771// CHECK3-LABEL: define {{[^@]+}}@_Z5tmainIiET_v772// CHECK3-SAME: () #[[ATTR4:[0-9]+]] comdat {773// CHECK3-NEXT: entry:774// CHECK3-NEXT: [[A:%.*]] = alloca [2 x i32], align 4775// CHECK3-NEXT: [[X_CASTED:%.*]] = alloca i32, align 4776// CHECK3-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [2 x ptr], align 4777// CHECK3-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [2 x ptr], align 4778// CHECK3-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [2 x ptr], align 4779// CHECK3-NEXT: [[TMP:%.*]] = alloca i32, align 4780// CHECK3-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8781// CHECK3-NEXT: [[TMP0:%.*]] = call align 4 ptr @llvm.threadlocal.address.p0(ptr align 4 @x)782// CHECK3-NEXT: [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4783// CHECK3-NEXT: store i32 [[TMP1]], ptr [[X_CASTED]], align 4784// CHECK3-NEXT: [[TMP2:%.*]] = load i32, ptr [[X_CASTED]], align 4785// CHECK3-NEXT: [[TMP3:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0786// CHECK3-NEXT: store i32 [[TMP2]], ptr [[TMP3]], align 4787// CHECK3-NEXT: [[TMP4:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0788// CHECK3-NEXT: store i32 [[TMP2]], ptr [[TMP4]], align 4789// CHECK3-NEXT: [[TMP5:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0790// CHECK3-NEXT: store ptr null, ptr [[TMP5]], align 4791// CHECK3-NEXT: [[TMP6:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1792// CHECK3-NEXT: store ptr [[A]], ptr [[TMP6]], align 4793// CHECK3-NEXT: [[TMP7:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1794// CHECK3-NEXT: store ptr [[A]], ptr [[TMP7]], align 4795// CHECK3-NEXT: [[TMP8:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 1796// CHECK3-NEXT: store ptr null, ptr [[TMP8]], align 4797// CHECK3-NEXT: [[TMP9:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0798// CHECK3-NEXT: [[TMP10:%.*]] = getelementptr inbounds [2 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0799// CHECK3-NEXT: [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0800// CHECK3-NEXT: store i32 3, ptr [[TMP11]], align 4801// CHECK3-NEXT: [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1802// CHECK3-NEXT: store i32 2, ptr [[TMP12]], align 4803// CHECK3-NEXT: [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2804// CHECK3-NEXT: store ptr [[TMP9]], ptr [[TMP13]], align 4805// CHECK3-NEXT: [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3806// CHECK3-NEXT: store ptr [[TMP10]], ptr [[TMP14]], align 4807// CHECK3-NEXT: [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4808// CHECK3-NEXT: store ptr @.offload_sizes.1, ptr [[TMP15]], align 4809// CHECK3-NEXT: [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5810// CHECK3-NEXT: store ptr @.offload_maptypes.2, ptr [[TMP16]], align 4811// CHECK3-NEXT: [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6812// CHECK3-NEXT: store ptr null, ptr [[TMP17]], align 4813// CHECK3-NEXT: [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7814// CHECK3-NEXT: store ptr null, ptr [[TMP18]], align 4815// CHECK3-NEXT: [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8816// CHECK3-NEXT: store i64 2, ptr [[TMP19]], align 8817// CHECK3-NEXT: [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9818// CHECK3-NEXT: store i64 0, ptr [[TMP20]], align 8819// CHECK3-NEXT: [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10820// CHECK3-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP21]], align 4821// CHECK3-NEXT: [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11822// CHECK3-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP22]], align 4823// CHECK3-NEXT: [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12824// CHECK3-NEXT: store i32 0, ptr [[TMP23]], align 4825// CHECK3-NEXT: [[TMP24:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB3]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.region_id, ptr [[KERNEL_ARGS]])826// CHECK3-NEXT: [[TMP25:%.*]] = icmp ne i32 [[TMP24]], 0827// CHECK3-NEXT: br i1 [[TMP25]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]828// CHECK3: omp_offload.failed:829// CHECK3-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34(i32 [[TMP2]], ptr [[A]]) #[[ATTR2]]830// CHECK3-NEXT: br label [[OMP_OFFLOAD_CONT]]831// CHECK3: omp_offload.cont:832// CHECK3-NEXT: ret i32 0833//834//835// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34836// CHECK3-SAME: (i32 noundef [[X:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]]) #[[ATTR1]] {837// CHECK3-NEXT: entry:838// CHECK3-NEXT: [[X_ADDR:%.*]] = alloca i32, align 4839// CHECK3-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 4840// CHECK3-NEXT: store i32 [[X]], ptr [[X_ADDR]], align 4841// CHECK3-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 4842// CHECK3-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 4843// CHECK3-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB3]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.omp_outlined, ptr [[TMP0]], ptr [[X_ADDR]])844// CHECK3-NEXT: ret void845//846//847// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.omp_outlined848// CHECK3-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR1]] {849// CHECK3-NEXT: entry:850// CHECK3-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4851// CHECK3-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4852// CHECK3-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 4853// CHECK3-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 4854// CHECK3-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4855// CHECK3-NEXT: [[TMP:%.*]] = alloca i32, align 4856// CHECK3-NEXT: [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4857// CHECK3-NEXT: [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4858// CHECK3-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4859// CHECK3-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4860// CHECK3-NEXT: [[I:%.*]] = alloca i32, align 4861// CHECK3-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4862// CHECK3-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4863// CHECK3-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 4864// CHECK3-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 4865// CHECK3-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 4866// CHECK3-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 4867// CHECK3-NEXT: store i32 0, ptr [[DOTOMP_COMB_LB]], align 4868// CHECK3-NEXT: store i32 1, ptr [[DOTOMP_COMB_UB]], align 4869// CHECK3-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4870// CHECK3-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4871// CHECK3-NEXT: [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4872// CHECK3-NEXT: [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 4873// CHECK3-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP3]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)874// CHECK3-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4875// CHECK3-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 1876// CHECK3-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]877// CHECK3: cond.true:878// CHECK3-NEXT: br label [[COND_END:%.*]]879// CHECK3: cond.false:880// CHECK3-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4881// CHECK3-NEXT: br label [[COND_END]]882// CHECK3: cond.end:883// CHECK3-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP5]], [[COND_FALSE]] ]884// CHECK3-NEXT: store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4885// CHECK3-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4886// CHECK3-NEXT: store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 4887// CHECK3-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]888// CHECK3: omp.inner.for.cond:889// CHECK3-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4890// CHECK3-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4891// CHECK3-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]892// CHECK3-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]893// CHECK3: omp.inner.for.body:894// CHECK3-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4895// CHECK3-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4896// CHECK3-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB3]], i32 4, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.omp_outlined.omp_outlined, i32 [[TMP9]], i32 [[TMP10]], ptr [[TMP0]], ptr [[TMP1]])897// CHECK3-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]898// CHECK3: omp.inner.for.inc:899// CHECK3-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4900// CHECK3-NEXT: [[TMP12:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4901// CHECK3-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP11]], [[TMP12]]902// CHECK3-NEXT: store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4903// CHECK3-NEXT: br label [[OMP_INNER_FOR_COND]]904// CHECK3: omp.inner.for.end:905// CHECK3-NEXT: br label [[OMP_LOOP_EXIT:%.*]]906// CHECK3: omp.loop.exit:907// CHECK3-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP3]])908// CHECK3-NEXT: ret void909//910//911// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l34.omp_outlined.omp_outlined912// CHECK3-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], i32 noundef [[DOTPREVIOUS_LB_:%.*]], i32 noundef [[DOTPREVIOUS_UB_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR1]] {913// CHECK3-NEXT: entry:914// CHECK3-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4915// CHECK3-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4916// CHECK3-NEXT: [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i32, align 4917// CHECK3-NEXT: [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i32, align 4918// CHECK3-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 4919// CHECK3-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 4920// CHECK3-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4921// CHECK3-NEXT: [[TMP:%.*]] = alloca i32, align 4922// CHECK3-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 4923// CHECK3-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 4924// CHECK3-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4925// CHECK3-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4926// CHECK3-NEXT: [[I:%.*]] = alloca i32, align 4927// CHECK3-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4928// CHECK3-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4929// CHECK3-NEXT: store i32 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 4930// CHECK3-NEXT: store i32 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 4931// CHECK3-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 4932// CHECK3-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 4933// CHECK3-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 4934// CHECK3-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 4935// CHECK3-NEXT: store i32 0, ptr [[DOTOMP_LB]], align 4936// CHECK3-NEXT: store i32 1, ptr [[DOTOMP_UB]], align 4937// CHECK3-NEXT: [[TMP2:%.*]] = load i32, ptr [[DOTPREVIOUS_LB__ADDR]], align 4938// CHECK3-NEXT: [[TMP3:%.*]] = load i32, ptr [[DOTPREVIOUS_UB__ADDR]], align 4939// CHECK3-NEXT: store i32 [[TMP2]], ptr [[DOTOMP_LB]], align 4940// CHECK3-NEXT: store i32 [[TMP3]], ptr [[DOTOMP_UB]], align 4941// CHECK3-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4942// CHECK3-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4943// CHECK3-NEXT: [[TMP4:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4944// CHECK3-NEXT: [[TMP5:%.*]] = load i32, ptr [[TMP4]], align 4945// CHECK3-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB2]], i32 [[TMP5]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)946// CHECK3-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4947// CHECK3-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP6]], 1948// CHECK3-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]949// CHECK3: cond.true:950// CHECK3-NEXT: br label [[COND_END:%.*]]951// CHECK3: cond.false:952// CHECK3-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4953// CHECK3-NEXT: br label [[COND_END]]954// CHECK3: cond.end:955// CHECK3-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP7]], [[COND_FALSE]] ]956// CHECK3-NEXT: store i32 [[COND]], ptr [[DOTOMP_UB]], align 4957// CHECK3-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4958// CHECK3-NEXT: store i32 [[TMP8]], ptr [[DOTOMP_IV]], align 4959// CHECK3-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]960// CHECK3: omp.inner.for.cond:961// CHECK3-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4962// CHECK3-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4963// CHECK3-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP9]], [[TMP10]]964// CHECK3-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]965// CHECK3: omp.inner.for.body:966// CHECK3-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4967// CHECK3-NEXT: [[MUL:%.*]] = mul nsw i32 [[TMP11]], 1968// CHECK3-NEXT: [[ADD:%.*]] = add nsw i32 0, [[MUL]]969// CHECK3-NEXT: store i32 [[ADD]], ptr [[I]], align 4970// CHECK3-NEXT: [[TMP12:%.*]] = load i32, ptr [[TMP1]], align 4971// CHECK3-NEXT: [[TMP13:%.*]] = load i32, ptr [[I]], align 4972// CHECK3-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[TMP0]], i32 0, i32 [[TMP13]]973// CHECK3-NEXT: store i32 [[TMP12]], ptr [[ARRAYIDX]], align 4974// CHECK3-NEXT: br label [[OMP_BODY_CONTINUE:%.*]]975// CHECK3: omp.body.continue:976// CHECK3-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]977// CHECK3: omp.inner.for.inc:978// CHECK3-NEXT: [[TMP14:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4979// CHECK3-NEXT: [[ADD2:%.*]] = add nsw i32 [[TMP14]], 1980// CHECK3-NEXT: store i32 [[ADD2]], ptr [[DOTOMP_IV]], align 4981// CHECK3-NEXT: br label [[OMP_INNER_FOR_COND]]982// CHECK3: omp.inner.for.end:983// CHECK3-NEXT: br label [[OMP_LOOP_EXIT:%.*]]984// CHECK3: omp.loop.exit:985// CHECK3-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP5]])986// CHECK3-NEXT: ret void987//988//989// CHECK3-LABEL: define {{[^@]+}}@_ZTW1x990// CHECK3-SAME: () #[[ATTR5:[0-9]+]] comdat {991// CHECK3-NEXT: [[TMP1:%.*]] = call align 4 ptr @llvm.threadlocal.address.p0(ptr align 4 @x)992// CHECK3-NEXT: ret ptr [[TMP1]]993//994//995// CHECK9-LABEL: define {{[^@]+}}@main996// CHECK9-SAME: () #[[ATTR0:[0-9]+]] {997// CHECK9-NEXT: entry:998// CHECK9-NEXT: [[RETVAL:%.*]] = alloca i32, align 4999// CHECK9-NEXT: [[A:%.*]] = alloca [2 x i32], align 41000// CHECK9-NEXT: [[REF_TMP:%.*]] = alloca [[CLASS_ANON:%.*]], align 81001// CHECK9-NEXT: store i32 0, ptr [[RETVAL]], align 41002// CHECK9-NEXT: [[TMP0:%.*]] = getelementptr inbounds nuw [[CLASS_ANON]], ptr [[REF_TMP]], i32 0, i32 01003// CHECK9-NEXT: store ptr [[A]], ptr [[TMP0]], align 81004// CHECK9-NEXT: call void @"_ZZ4mainENK3$_0clEv"(ptr noundef nonnull align 8 dereferenceable(8) [[REF_TMP]])1005// CHECK9-NEXT: ret i32 01006//1007//1008// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l461009// CHECK9-SAME: (i64 noundef [[X:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]]) #[[ATTR2:[0-9]+]] {1010// CHECK9-NEXT: entry:1011// CHECK9-NEXT: [[X_ADDR:%.*]] = alloca i64, align 81012// CHECK9-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 81013// CHECK9-NEXT: store i64 [[X]], ptr [[X_ADDR]], align 81014// CHECK9-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 81015// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 81016// CHECK9-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB3:[0-9]+]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l46.omp_outlined, ptr [[TMP0]], ptr [[X_ADDR]])1017// CHECK9-NEXT: ret void1018//1019//1020// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l46.omp_outlined1021// CHECK9-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR2]] {1022// CHECK9-NEXT: entry:1023// CHECK9-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 81024// CHECK9-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 81025// CHECK9-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 81026// CHECK9-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 81027// CHECK9-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 41028// CHECK9-NEXT: [[TMP:%.*]] = alloca i32, align 41029// CHECK9-NEXT: [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 41030// CHECK9-NEXT: [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 41031// CHECK9-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 41032// CHECK9-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 41033// CHECK9-NEXT: [[I:%.*]] = alloca i32, align 41034// CHECK9-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 81035// CHECK9-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 81036// CHECK9-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 81037// CHECK9-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 81038// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 81039// CHECK9-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 81040// CHECK9-NEXT: store i32 0, ptr [[DOTOMP_COMB_LB]], align 41041// CHECK9-NEXT: store i32 1, ptr [[DOTOMP_COMB_UB]], align 41042// CHECK9-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 41043// CHECK9-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 41044// CHECK9-NEXT: [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 81045// CHECK9-NEXT: [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 41046// CHECK9-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP3]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)1047// CHECK9-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41048// CHECK9-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 11049// CHECK9-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]1050// CHECK9: cond.true:1051// CHECK9-NEXT: br label [[COND_END:%.*]]1052// CHECK9: cond.false:1053// CHECK9-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41054// CHECK9-NEXT: br label [[COND_END]]1055// CHECK9: cond.end:1056// CHECK9-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP5]], [[COND_FALSE]] ]1057// CHECK9-NEXT: store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 41058// CHECK9-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 41059// CHECK9-NEXT: store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 41060// CHECK9-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]1061// CHECK9: omp.inner.for.cond:1062// CHECK9-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41063// CHECK9-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41064// CHECK9-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]1065// CHECK9-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]1066// CHECK9: omp.inner.for.body:1067// CHECK9-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 41068// CHECK9-NEXT: [[TMP10:%.*]] = zext i32 [[TMP9]] to i641069// CHECK9-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41070// CHECK9-NEXT: [[TMP12:%.*]] = zext i32 [[TMP11]] to i641071// CHECK9-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB3]], i32 4, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l46.omp_outlined.omp_outlined, i64 [[TMP10]], i64 [[TMP12]], ptr [[TMP0]], ptr [[TMP1]])1072// CHECK9-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]1073// CHECK9: omp.inner.for.inc:1074// CHECK9-NEXT: [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41075// CHECK9-NEXT: [[TMP14:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 41076// CHECK9-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP13]], [[TMP14]]1077// CHECK9-NEXT: store i32 [[ADD]], ptr [[DOTOMP_IV]], align 41078// CHECK9-NEXT: br label [[OMP_INNER_FOR_COND]]1079// CHECK9: omp.inner.for.end:1080// CHECK9-NEXT: br label [[OMP_LOOP_EXIT:%.*]]1081// CHECK9: omp.loop.exit:1082// CHECK9-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP3]])1083// CHECK9-NEXT: ret void1084//1085//1086// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l46.omp_outlined.omp_outlined1087// CHECK9-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], i64 noundef [[DOTPREVIOUS_LB_:%.*]], i64 noundef [[DOTPREVIOUS_UB_:%.*]], ptr noundef nonnull align 4 dereferenceable(8) [[A:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[X:%.*]]) #[[ATTR2]] {1088// CHECK9-NEXT: entry:1089// CHECK9-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 81090// CHECK9-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 81091// CHECK9-NEXT: [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 81092// CHECK9-NEXT: [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 81093// CHECK9-NEXT: [[A_ADDR:%.*]] = alloca ptr, align 81094// CHECK9-NEXT: [[X_ADDR:%.*]] = alloca ptr, align 81095// CHECK9-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 41096// CHECK9-NEXT: [[TMP:%.*]] = alloca i32, align 41097// CHECK9-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 41098// CHECK9-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 41099// CHECK9-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 41100// CHECK9-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 41101// CHECK9-NEXT: [[I:%.*]] = alloca i32, align 41102// CHECK9-NEXT: [[REF_TMP:%.*]] = alloca [[CLASS_ANON_0:%.*]], align 81103// CHECK9-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 81104// CHECK9-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 81105// CHECK9-NEXT: store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 81106// CHECK9-NEXT: store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 81107// CHECK9-NEXT: store ptr [[A]], ptr [[A_ADDR]], align 81108// CHECK9-NEXT: store ptr [[X]], ptr [[X_ADDR]], align 81109// CHECK9-NEXT: [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 81110// CHECK9-NEXT: [[TMP1:%.*]] = load ptr, ptr [[X_ADDR]], align 81111// CHECK9-NEXT: store i32 0, ptr [[DOTOMP_LB]], align 41112// CHECK9-NEXT: store i32 1, ptr [[DOTOMP_UB]], align 41113// CHECK9-NEXT: [[TMP2:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 81114// CHECK9-NEXT: [[CONV:%.*]] = trunc i64 [[TMP2]] to i321115// CHECK9-NEXT: [[TMP3:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 81116// CHECK9-NEXT: [[CONV1:%.*]] = trunc i64 [[TMP3]] to i321117// CHECK9-NEXT: store i32 [[CONV]], ptr [[DOTOMP_LB]], align 41118// CHECK9-NEXT: store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 41119// CHECK9-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 41120// CHECK9-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 41121// CHECK9-NEXT: [[TMP4:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 81122// CHECK9-NEXT: [[TMP5:%.*]] = load i32, ptr [[TMP4]], align 41123// CHECK9-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB2:[0-9]+]], i32 [[TMP5]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)1124// CHECK9-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41125// CHECK9-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP6]], 11126// CHECK9-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]1127// CHECK9: cond.true:1128// CHECK9-NEXT: br label [[COND_END:%.*]]1129// CHECK9: cond.false:1130// CHECK9-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41131// CHECK9-NEXT: br label [[COND_END]]1132// CHECK9: cond.end:1133// CHECK9-NEXT: [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP7]], [[COND_FALSE]] ]1134// CHECK9-NEXT: store i32 [[COND]], ptr [[DOTOMP_UB]], align 41135// CHECK9-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_LB]], align 41136// CHECK9-NEXT: store i32 [[TMP8]], ptr [[DOTOMP_IV]], align 41137// CHECK9-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]1138// CHECK9: omp.inner.for.cond:1139// CHECK9-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41140// CHECK9-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41141// CHECK9-NEXT: [[CMP2:%.*]] = icmp sle i32 [[TMP9]], [[TMP10]]1142// CHECK9-NEXT: br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]1143// CHECK9: omp.inner.for.body:1144// CHECK9-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41145// CHECK9-NEXT: [[MUL:%.*]] = mul nsw i32 [[TMP11]], 11146// CHECK9-NEXT: [[ADD:%.*]] = add nsw i32 0, [[MUL]]1147// CHECK9-NEXT: store i32 [[ADD]], ptr [[I]], align 41148// CHECK9-NEXT: [[TMP12:%.*]] = load i32, ptr [[TMP1]], align 41149// CHECK9-NEXT: [[TMP13:%.*]] = load i32, ptr [[I]], align 41150// CHECK9-NEXT: [[IDXPROM:%.*]] = sext i32 [[TMP13]] to i641151// CHECK9-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[TMP0]], i64 0, i64 [[IDXPROM]]1152// CHECK9-NEXT: store i32 [[TMP12]], ptr [[ARRAYIDX]], align 41153// CHECK9-NEXT: [[TMP14:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 01154// CHECK9-NEXT: store ptr [[TMP0]], ptr [[TMP14]], align 81155// CHECK9-NEXT: [[TMP15:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 11156// CHECK9-NEXT: store ptr [[I]], ptr [[TMP15]], align 81157// CHECK9-NEXT: [[TMP16:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 21158// CHECK9-NEXT: store ptr [[TMP1]], ptr [[TMP16]], align 81159// CHECK9-NEXT: call void @"_ZZZ4mainENK3$_0clEvENKUlvE_clEv"(ptr noundef nonnull align 8 dereferenceable(24) [[REF_TMP]])1160// CHECK9-NEXT: br label [[OMP_BODY_CONTINUE:%.*]]1161// CHECK9: omp.body.continue:1162// CHECK9-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]1163// CHECK9: omp.inner.for.inc:1164// CHECK9-NEXT: [[TMP17:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41165// CHECK9-NEXT: [[ADD3:%.*]] = add nsw i32 [[TMP17]], 11166// CHECK9-NEXT: store i32 [[ADD3]], ptr [[DOTOMP_IV]], align 41167// CHECK9-NEXT: br label [[OMP_INNER_FOR_COND]]1168// CHECK9: omp.inner.for.end:1169// CHECK9-NEXT: br label [[OMP_LOOP_EXIT:%.*]]1170// CHECK9: omp.loop.exit:1171// CHECK9-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP5]])1172// CHECK9-NEXT: ret void1173//1174//1175// CHECK9-LABEL: define {{[^@]+}}@_ZTW1x1176// CHECK9-SAME: () #[[ATTR5:[0-9]+]] comdat {1177// CHECK9-NEXT: [[TMP1:%.*]] = call align 4 ptr @llvm.threadlocal.address.p0(ptr align 4 @x)1178// CHECK9-NEXT: ret ptr [[TMP1]]1179//1180