brintos / llvm-project-archived public Read only

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 28template <typename T>29T tmain() {30  T t_var = T();31  T vec[] = {1, 2};32#pragma omp target33#pragma omp teams distribute parallel for reduction(+: t_var)34  for (int i = 0; i < 2; ++i) {35    t_var += (T) i;36  }37  return T();38}39 40int main() {41  static int sivar;42#ifdef LAMBDA43 44  [&]() {45#pragma omp target46#pragma omp teams distribute parallel for reduction(+: sivar)47  for (int i = 0; i < 2; ++i) {48 49    // Skip global and bound tid vars50 51 52 53    // Skip global and bound tid vars, and prev lb and ub vars54    // skip loop vars55 56 57    sivar += i;58 59    [&]() {60 61      sivar += 4;62 63    }();64  }65  }();66  return 0;67#else68#pragma omp target69#pragma omp teams distribute parallel for reduction(+: sivar)70  for (int i = 0; i < 2; ++i) {71    sivar += i;72  }73  return tmain<int>();74#endif75}76 77 78 79 80// Skip global and bound tid vars81 82 83// Skip global and bound tid vars, and prev lb and ub84// skip loop vars85 86 87 88 89// Skip global and bound tid vars90 91 92// Skip global and bound tid vars, and prev lb and ub vars93// skip loop vars94 95#endif96// CHECK1-LABEL: define {{[^@]+}}@main97// CHECK1-SAME: () #[[ATTR0:[0-9]+]] {98// CHECK1-NEXT:  entry:99// CHECK1-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4100// CHECK1-NEXT:    [[SIVAR_CASTED:%.*]] = alloca i64, align 8101// CHECK1-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 8102// CHECK1-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 8103// CHECK1-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 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:%.*]] = load i32, ptr @_ZZ4mainE5sivar, align 4108// CHECK1-NEXT:    store i32 [[TMP0]], ptr [[SIVAR_CASTED]], align 4109// CHECK1-NEXT:    [[TMP1:%.*]] = load i64, ptr [[SIVAR_CASTED]], align 8110// CHECK1-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0111// CHECK1-NEXT:    store i64 [[TMP1]], ptr [[TMP2]], align 8112// CHECK1-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0113// CHECK1-NEXT:    store i64 [[TMP1]], ptr [[TMP3]], align 8114// CHECK1-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0115// CHECK1-NEXT:    store ptr null, ptr [[TMP4]], align 8116// CHECK1-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0117// CHECK1-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0118// CHECK1-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0119// CHECK1-NEXT:    store i32 3, ptr [[TMP7]], align 4120// CHECK1-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1121// CHECK1-NEXT:    store i32 1, ptr [[TMP8]], align 4122// CHECK1-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2123// CHECK1-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 8124// CHECK1-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3125// CHECK1-NEXT:    store ptr [[TMP6]], ptr [[TMP10]], align 8126// CHECK1-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4127// CHECK1-NEXT:    store ptr @.offload_sizes, ptr [[TMP11]], align 8128// CHECK1-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5129// CHECK1-NEXT:    store ptr @.offload_maptypes, ptr [[TMP12]], align 8130// CHECK1-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6131// CHECK1-NEXT:    store ptr null, ptr [[TMP13]], align 8132// CHECK1-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7133// CHECK1-NEXT:    store ptr null, ptr [[TMP14]], align 8134// CHECK1-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8135// CHECK1-NEXT:    store i64 2, ptr [[TMP15]], align 8136// CHECK1-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9137// CHECK1-NEXT:    store i64 0, ptr [[TMP16]], align 8138// CHECK1-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10139// CHECK1-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP17]], align 4140// CHECK1-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11141// CHECK1-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP18]], align 4142// CHECK1-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12143// CHECK1-NEXT:    store i32 0, ptr [[TMP19]], align 4144// CHECK1-NEXT:    [[TMP20:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB4:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.region_id, ptr [[KERNEL_ARGS]])145// CHECK1-NEXT:    [[TMP21:%.*]] = icmp ne i32 [[TMP20]], 0146// CHECK1-NEXT:    br i1 [[TMP21]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]147// CHECK1:       omp_offload.failed:148// CHECK1-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68(i64 [[TMP1]]) #[[ATTR2:[0-9]+]]149// CHECK1-NEXT:    br label [[OMP_OFFLOAD_CONT]]150// CHECK1:       omp_offload.cont:151// CHECK1-NEXT:    [[CALL:%.*]] = call noundef signext i32 @_Z5tmainIiET_v()152// CHECK1-NEXT:    ret i32 [[CALL]]153//154//155// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68156// CHECK1-SAME: (i64 noundef [[SIVAR:%.*]]) #[[ATTR1:[0-9]+]] {157// CHECK1-NEXT:  entry:158// CHECK1-NEXT:    [[SIVAR_ADDR:%.*]] = alloca i64, align 8159// CHECK1-NEXT:    store i64 [[SIVAR]], ptr [[SIVAR_ADDR]], align 8160// CHECK1-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB4]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined, ptr [[SIVAR_ADDR]])161// CHECK1-NEXT:    ret void162//163//164// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined165// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[SIVAR:%.*]]) #[[ATTR1]] {166// CHECK1-NEXT:  entry:167// CHECK1-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8168// CHECK1-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8169// CHECK1-NEXT:    [[SIVAR_ADDR:%.*]] = alloca ptr, align 8170// CHECK1-NEXT:    [[SIVAR1:%.*]] = alloca i32, align 4171// CHECK1-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4172// CHECK1-NEXT:    [[TMP:%.*]] = alloca i32, align 4173// CHECK1-NEXT:    [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4174// CHECK1-NEXT:    [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4175// CHECK1-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4176// CHECK1-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4177// CHECK1-NEXT:    [[I:%.*]] = alloca i32, align 4178// CHECK1-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 8179// CHECK1-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8180// CHECK1-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8181// CHECK1-NEXT:    store ptr [[SIVAR]], ptr [[SIVAR_ADDR]], align 8182// CHECK1-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[SIVAR_ADDR]], align 8183// CHECK1-NEXT:    store i32 0, ptr [[SIVAR1]], align 4184// CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_COMB_LB]], align 4185// CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_COMB_UB]], align 4186// CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4187// CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4188// CHECK1-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8189// CHECK1-NEXT:    [[TMP2:%.*]] = load i32, ptr [[TMP1]], align 4190// CHECK1-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP2]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)191// CHECK1-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4192// CHECK1-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP3]], 1193// CHECK1-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]194// CHECK1:       cond.true:195// CHECK1-NEXT:    br label [[COND_END:%.*]]196// CHECK1:       cond.false:197// CHECK1-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4198// CHECK1-NEXT:    br label [[COND_END]]199// CHECK1:       cond.end:200// CHECK1-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP4]], [[COND_FALSE]] ]201// CHECK1-NEXT:    store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4202// CHECK1-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4203// CHECK1-NEXT:    store i32 [[TMP5]], ptr [[DOTOMP_IV]], align 4204// CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]205// CHECK1:       omp.inner.for.cond:206// CHECK1-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4207// CHECK1-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4208// CHECK1-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP6]], [[TMP7]]209// CHECK1-NEXT:    br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]210// CHECK1:       omp.inner.for.body:211// CHECK1-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4212// CHECK1-NEXT:    [[TMP9:%.*]] = zext i32 [[TMP8]] to i64213// CHECK1-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4214// CHECK1-NEXT:    [[TMP11:%.*]] = zext i32 [[TMP10]] to i64215// CHECK1-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB4]], i32 3, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp_outlined, i64 [[TMP9]], i64 [[TMP11]], ptr [[SIVAR1]])216// CHECK1-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]217// CHECK1:       omp.inner.for.inc:218// CHECK1-NEXT:    [[TMP12:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4219// CHECK1-NEXT:    [[TMP13:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4220// CHECK1-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP12]], [[TMP13]]221// CHECK1-NEXT:    store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4222// CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND]]223// CHECK1:       omp.inner.for.end:224// CHECK1-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]225// CHECK1:       omp.loop.exit:226// CHECK1-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP2]])227// CHECK1-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i64 0, i64 0228// CHECK1-NEXT:    store ptr [[SIVAR1]], ptr [[TMP14]], align 8229// CHECK1-NEXT:    [[TMP15:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3:[0-9]+]], i32 [[TMP2]], i32 1, i64 8, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)230// CHECK1-NEXT:    switch i32 [[TMP15]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [231// CHECK1-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]232// CHECK1-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]233// CHECK1-NEXT:    ]234// CHECK1:       .omp.reduction.case1:235// CHECK1-NEXT:    [[TMP16:%.*]] = load i32, ptr [[TMP0]], align 4236// CHECK1-NEXT:    [[TMP17:%.*]] = load i32, ptr [[SIVAR1]], align 4237// CHECK1-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP16]], [[TMP17]]238// CHECK1-NEXT:    store i32 [[ADD3]], ptr [[TMP0]], align 4239// CHECK1-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP2]], ptr @.gomp_critical_user_.reduction.var)240// CHECK1-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]241// CHECK1:       .omp.reduction.case2:242// CHECK1-NEXT:    [[TMP18:%.*]] = load i32, ptr [[SIVAR1]], align 4243// CHECK1-NEXT:    [[TMP19:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP18]] monotonic, align 4244// CHECK1-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]245// CHECK1:       .omp.reduction.default:246// CHECK1-NEXT:    ret void247//248//249// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp_outlined250// 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(4) [[SIVAR:%.*]]) #[[ATTR1]] {251// CHECK1-NEXT:  entry:252// CHECK1-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8253// CHECK1-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8254// CHECK1-NEXT:    [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 8255// CHECK1-NEXT:    [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 8256// CHECK1-NEXT:    [[SIVAR_ADDR:%.*]] = alloca ptr, align 8257// CHECK1-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4258// CHECK1-NEXT:    [[TMP:%.*]] = alloca i32, align 4259// CHECK1-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4260// CHECK1-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4261// CHECK1-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4262// CHECK1-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4263// CHECK1-NEXT:    [[SIVAR2:%.*]] = alloca i32, align 4264// CHECK1-NEXT:    [[I:%.*]] = alloca i32, align 4265// CHECK1-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 8266// CHECK1-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8267// CHECK1-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8268// CHECK1-NEXT:    store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 8269// CHECK1-NEXT:    store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 8270// CHECK1-NEXT:    store ptr [[SIVAR]], ptr [[SIVAR_ADDR]], align 8271// CHECK1-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[SIVAR_ADDR]], align 8272// CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4273// CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4274// CHECK1-NEXT:    [[TMP1:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 8275// CHECK1-NEXT:    [[CONV:%.*]] = trunc i64 [[TMP1]] to i32276// CHECK1-NEXT:    [[TMP2:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 8277// CHECK1-NEXT:    [[CONV1:%.*]] = trunc i64 [[TMP2]] to i32278// CHECK1-NEXT:    store i32 [[CONV]], ptr [[DOTOMP_LB]], align 4279// CHECK1-NEXT:    store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 4280// CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4281// CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4282// CHECK1-NEXT:    store i32 0, ptr [[SIVAR2]], align 4283// CHECK1-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8284// CHECK1-NEXT:    [[TMP4:%.*]] = load i32, ptr [[TMP3]], align 4285// CHECK1-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB2:[0-9]+]], i32 [[TMP4]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)286// CHECK1-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4287// CHECK1-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP5]], 1288// CHECK1-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]289// CHECK1:       cond.true:290// CHECK1-NEXT:    br label [[COND_END:%.*]]291// CHECK1:       cond.false:292// CHECK1-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4293// CHECK1-NEXT:    br label [[COND_END]]294// CHECK1:       cond.end:295// CHECK1-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP6]], [[COND_FALSE]] ]296// CHECK1-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4297// CHECK1-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4298// CHECK1-NEXT:    store i32 [[TMP7]], ptr [[DOTOMP_IV]], align 4299// CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]300// CHECK1:       omp.inner.for.cond:301// CHECK1-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4302// CHECK1-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4303// CHECK1-NEXT:    [[CMP3:%.*]] = icmp sle i32 [[TMP8]], [[TMP9]]304// CHECK1-NEXT:    br i1 [[CMP3]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]305// CHECK1:       omp.inner.for.body:306// CHECK1-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4307// CHECK1-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP10]], 1308// CHECK1-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]309// CHECK1-NEXT:    store i32 [[ADD]], ptr [[I]], align 4310// CHECK1-NEXT:    [[TMP11:%.*]] = load i32, ptr [[I]], align 4311// CHECK1-NEXT:    [[TMP12:%.*]] = load i32, ptr [[SIVAR2]], align 4312// CHECK1-NEXT:    [[ADD4:%.*]] = add nsw i32 [[TMP12]], [[TMP11]]313// CHECK1-NEXT:    store i32 [[ADD4]], ptr [[SIVAR2]], align 4314// CHECK1-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]315// CHECK1:       omp.body.continue:316// CHECK1-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]317// CHECK1:       omp.inner.for.inc:318// CHECK1-NEXT:    [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4319// CHECK1-NEXT:    [[ADD5:%.*]] = add nsw i32 [[TMP13]], 1320// CHECK1-NEXT:    store i32 [[ADD5]], ptr [[DOTOMP_IV]], align 4321// CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND]]322// CHECK1:       omp.inner.for.end:323// CHECK1-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]324// CHECK1:       omp.loop.exit:325// CHECK1-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP4]])326// CHECK1-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i64 0, i64 0327// CHECK1-NEXT:    store ptr [[SIVAR2]], ptr [[TMP14]], align 8328// CHECK1-NEXT:    [[TMP15:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], i32 1, i64 8, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)329// CHECK1-NEXT:    switch i32 [[TMP15]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [330// CHECK1-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]331// CHECK1-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]332// CHECK1-NEXT:    ]333// CHECK1:       .omp.reduction.case1:334// CHECK1-NEXT:    [[TMP16:%.*]] = load i32, ptr [[TMP0]], align 4335// CHECK1-NEXT:    [[TMP17:%.*]] = load i32, ptr [[SIVAR2]], align 4336// CHECK1-NEXT:    [[ADD6:%.*]] = add nsw i32 [[TMP16]], [[TMP17]]337// CHECK1-NEXT:    store i32 [[ADD6]], ptr [[TMP0]], align 4338// CHECK1-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], ptr @.gomp_critical_user_.reduction.var)339// CHECK1-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]340// CHECK1:       .omp.reduction.case2:341// CHECK1-NEXT:    [[TMP18:%.*]] = load i32, ptr [[SIVAR2]], align 4342// CHECK1-NEXT:    [[TMP19:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP18]] monotonic, align 4343// CHECK1-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]344// CHECK1:       .omp.reduction.default:345// CHECK1-NEXT:    ret void346//347//348// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp_outlined.omp.reduction.reduction_func349// CHECK1-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR3:[0-9]+]] {350// CHECK1-NEXT:  entry:351// CHECK1-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 8352// CHECK1-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 8353// CHECK1-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 8354// CHECK1-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 8355// CHECK1-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 8356// CHECK1-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 8357// CHECK1-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i64 0, i64 0358// CHECK1-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 8359// CHECK1-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i64 0, i64 0360// CHECK1-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 8361// CHECK1-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 4362// CHECK1-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 4363// CHECK1-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]364// CHECK1-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 4365// CHECK1-NEXT:    ret void366//367//368// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp.reduction.reduction_func369// CHECK1-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR3]] {370// CHECK1-NEXT:  entry:371// CHECK1-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 8372// CHECK1-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 8373// CHECK1-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 8374// CHECK1-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 8375// CHECK1-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 8376// CHECK1-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 8377// CHECK1-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i64 0, i64 0378// CHECK1-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 8379// CHECK1-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i64 0, i64 0380// CHECK1-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 8381// CHECK1-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 4382// CHECK1-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 4383// CHECK1-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]384// CHECK1-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 4385// CHECK1-NEXT:    ret void386//387//388// CHECK1-LABEL: define {{[^@]+}}@_Z5tmainIiET_v389// CHECK1-SAME: () #[[ATTR5:[0-9]+]] comdat {390// CHECK1-NEXT:  entry:391// CHECK1-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4392// CHECK1-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4393// CHECK1-NEXT:    [[T_VAR_CASTED:%.*]] = alloca i64, align 8394// CHECK1-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 8395// CHECK1-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 8396// CHECK1-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 8397// CHECK1-NEXT:    [[TMP:%.*]] = alloca i32, align 4398// CHECK1-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8399// CHECK1-NEXT:    store i32 0, ptr [[T_VAR]], align 4400// CHECK1-NEXT:    call void @llvm.memcpy.p0.p0.i64(ptr align 4 [[VEC]], ptr align 4 @__const._Z5tmainIiET_v.vec, i64 8, i1 false)401// CHECK1-NEXT:    [[TMP0:%.*]] = load i32, ptr [[T_VAR]], align 4402// CHECK1-NEXT:    store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 4403// CHECK1-NEXT:    [[TMP1:%.*]] = load i64, ptr [[T_VAR_CASTED]], align 8404// CHECK1-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0405// CHECK1-NEXT:    store i64 [[TMP1]], ptr [[TMP2]], align 8406// CHECK1-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0407// CHECK1-NEXT:    store i64 [[TMP1]], ptr [[TMP3]], align 8408// CHECK1-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0409// CHECK1-NEXT:    store ptr null, ptr [[TMP4]], align 8410// CHECK1-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0411// CHECK1-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0412// CHECK1-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0413// CHECK1-NEXT:    store i32 3, ptr [[TMP7]], align 4414// CHECK1-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1415// CHECK1-NEXT:    store i32 1, ptr [[TMP8]], align 4416// CHECK1-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2417// CHECK1-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 8418// CHECK1-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3419// CHECK1-NEXT:    store ptr [[TMP6]], ptr [[TMP10]], align 8420// CHECK1-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4421// CHECK1-NEXT:    store ptr @.offload_sizes.1, ptr [[TMP11]], align 8422// CHECK1-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5423// CHECK1-NEXT:    store ptr @.offload_maptypes.2, ptr [[TMP12]], align 8424// CHECK1-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6425// CHECK1-NEXT:    store ptr null, ptr [[TMP13]], align 8426// CHECK1-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7427// CHECK1-NEXT:    store ptr null, ptr [[TMP14]], align 8428// CHECK1-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8429// CHECK1-NEXT:    store i64 2, ptr [[TMP15]], align 8430// CHECK1-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9431// CHECK1-NEXT:    store i64 0, ptr [[TMP16]], align 8432// CHECK1-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10433// CHECK1-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP17]], align 4434// CHECK1-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11435// CHECK1-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP18]], align 4436// CHECK1-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12437// CHECK1-NEXT:    store i32 0, ptr [[TMP19]], align 4438// CHECK1-NEXT:    [[TMP20:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB4]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.region_id, ptr [[KERNEL_ARGS]])439// CHECK1-NEXT:    [[TMP21:%.*]] = icmp ne i32 [[TMP20]], 0440// CHECK1-NEXT:    br i1 [[TMP21]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]441// CHECK1:       omp_offload.failed:442// CHECK1-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32(i64 [[TMP1]]) #[[ATTR2]]443// CHECK1-NEXT:    br label [[OMP_OFFLOAD_CONT]]444// CHECK1:       omp_offload.cont:445// CHECK1-NEXT:    ret i32 0446//447//448// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32449// CHECK1-SAME: (i64 noundef [[T_VAR:%.*]]) #[[ATTR1]] {450// CHECK1-NEXT:  entry:451// CHECK1-NEXT:    [[T_VAR_ADDR:%.*]] = alloca i64, align 8452// CHECK1-NEXT:    store i64 [[T_VAR]], ptr [[T_VAR_ADDR]], align 8453// CHECK1-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB4]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined, ptr [[T_VAR_ADDR]])454// CHECK1-NEXT:    ret void455//456//457// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined458// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[T_VAR:%.*]]) #[[ATTR1]] {459// CHECK1-NEXT:  entry:460// CHECK1-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8461// CHECK1-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8462// CHECK1-NEXT:    [[T_VAR_ADDR:%.*]] = alloca ptr, align 8463// CHECK1-NEXT:    [[T_VAR1:%.*]] = alloca i32, align 4464// CHECK1-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4465// CHECK1-NEXT:    [[TMP:%.*]] = alloca i32, align 4466// CHECK1-NEXT:    [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4467// CHECK1-NEXT:    [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4468// CHECK1-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4469// CHECK1-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4470// CHECK1-NEXT:    [[I:%.*]] = alloca i32, align 4471// CHECK1-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 8472// CHECK1-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8473// CHECK1-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8474// CHECK1-NEXT:    store ptr [[T_VAR]], ptr [[T_VAR_ADDR]], align 8475// CHECK1-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[T_VAR_ADDR]], align 8476// CHECK1-NEXT:    store i32 0, ptr [[T_VAR1]], align 4477// CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_COMB_LB]], align 4478// CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_COMB_UB]], align 4479// CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4480// CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4481// CHECK1-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8482// CHECK1-NEXT:    [[TMP2:%.*]] = load i32, ptr [[TMP1]], align 4483// CHECK1-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP2]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)484// CHECK1-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4485// CHECK1-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP3]], 1486// CHECK1-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]487// CHECK1:       cond.true:488// CHECK1-NEXT:    br label [[COND_END:%.*]]489// CHECK1:       cond.false:490// CHECK1-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4491// CHECK1-NEXT:    br label [[COND_END]]492// CHECK1:       cond.end:493// CHECK1-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP4]], [[COND_FALSE]] ]494// CHECK1-NEXT:    store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4495// CHECK1-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4496// CHECK1-NEXT:    store i32 [[TMP5]], ptr [[DOTOMP_IV]], align 4497// CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]498// CHECK1:       omp.inner.for.cond:499// CHECK1-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4500// CHECK1-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4501// CHECK1-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP6]], [[TMP7]]502// CHECK1-NEXT:    br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]503// CHECK1:       omp.inner.for.body:504// CHECK1-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4505// CHECK1-NEXT:    [[TMP9:%.*]] = zext i32 [[TMP8]] to i64506// CHECK1-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4507// CHECK1-NEXT:    [[TMP11:%.*]] = zext i32 [[TMP10]] to i64508// CHECK1-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB4]], i32 3, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp_outlined, i64 [[TMP9]], i64 [[TMP11]], ptr [[T_VAR1]])509// CHECK1-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]510// CHECK1:       omp.inner.for.inc:511// CHECK1-NEXT:    [[TMP12:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4512// CHECK1-NEXT:    [[TMP13:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4513// CHECK1-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP12]], [[TMP13]]514// CHECK1-NEXT:    store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4515// CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND]]516// CHECK1:       omp.inner.for.end:517// CHECK1-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]518// CHECK1:       omp.loop.exit:519// CHECK1-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP2]])520// CHECK1-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i64 0, i64 0521// CHECK1-NEXT:    store ptr [[T_VAR1]], ptr [[TMP14]], align 8522// CHECK1-NEXT:    [[TMP15:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP2]], i32 1, i64 8, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)523// CHECK1-NEXT:    switch i32 [[TMP15]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [524// CHECK1-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]525// CHECK1-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]526// CHECK1-NEXT:    ]527// CHECK1:       .omp.reduction.case1:528// CHECK1-NEXT:    [[TMP16:%.*]] = load i32, ptr [[TMP0]], align 4529// CHECK1-NEXT:    [[TMP17:%.*]] = load i32, ptr [[T_VAR1]], align 4530// CHECK1-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP16]], [[TMP17]]531// CHECK1-NEXT:    store i32 [[ADD3]], ptr [[TMP0]], align 4532// CHECK1-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP2]], ptr @.gomp_critical_user_.reduction.var)533// CHECK1-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]534// CHECK1:       .omp.reduction.case2:535// CHECK1-NEXT:    [[TMP18:%.*]] = load i32, ptr [[T_VAR1]], align 4536// CHECK1-NEXT:    [[TMP19:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP18]] monotonic, align 4537// CHECK1-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]538// CHECK1:       .omp.reduction.default:539// CHECK1-NEXT:    ret void540//541//542// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp_outlined543// 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(4) [[T_VAR:%.*]]) #[[ATTR1]] {544// CHECK1-NEXT:  entry:545// CHECK1-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8546// CHECK1-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8547// CHECK1-NEXT:    [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 8548// CHECK1-NEXT:    [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 8549// CHECK1-NEXT:    [[T_VAR_ADDR:%.*]] = alloca ptr, align 8550// CHECK1-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4551// CHECK1-NEXT:    [[TMP:%.*]] = alloca i32, align 4552// CHECK1-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4553// CHECK1-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4554// CHECK1-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4555// CHECK1-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4556// CHECK1-NEXT:    [[T_VAR2:%.*]] = alloca i32, align 4557// CHECK1-NEXT:    [[I:%.*]] = alloca i32, align 4558// CHECK1-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 8559// CHECK1-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8560// CHECK1-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8561// CHECK1-NEXT:    store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 8562// CHECK1-NEXT:    store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 8563// CHECK1-NEXT:    store ptr [[T_VAR]], ptr [[T_VAR_ADDR]], align 8564// CHECK1-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[T_VAR_ADDR]], align 8565// CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4566// CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4567// CHECK1-NEXT:    [[TMP1:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 8568// CHECK1-NEXT:    [[CONV:%.*]] = trunc i64 [[TMP1]] to i32569// CHECK1-NEXT:    [[TMP2:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 8570// CHECK1-NEXT:    [[CONV1:%.*]] = trunc i64 [[TMP2]] to i32571// CHECK1-NEXT:    store i32 [[CONV]], ptr [[DOTOMP_LB]], align 4572// CHECK1-NEXT:    store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 4573// CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4574// CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4575// CHECK1-NEXT:    store i32 0, ptr [[T_VAR2]], align 4576// CHECK1-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8577// CHECK1-NEXT:    [[TMP4:%.*]] = load i32, ptr [[TMP3]], align 4578// CHECK1-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB2]], i32 [[TMP4]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)579// CHECK1-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4580// CHECK1-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP5]], 1581// CHECK1-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]582// CHECK1:       cond.true:583// CHECK1-NEXT:    br label [[COND_END:%.*]]584// CHECK1:       cond.false:585// CHECK1-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4586// CHECK1-NEXT:    br label [[COND_END]]587// CHECK1:       cond.end:588// CHECK1-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP6]], [[COND_FALSE]] ]589// CHECK1-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4590// CHECK1-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4591// CHECK1-NEXT:    store i32 [[TMP7]], ptr [[DOTOMP_IV]], align 4592// CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]593// CHECK1:       omp.inner.for.cond:594// CHECK1-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4595// CHECK1-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4596// CHECK1-NEXT:    [[CMP3:%.*]] = icmp sle i32 [[TMP8]], [[TMP9]]597// CHECK1-NEXT:    br i1 [[CMP3]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]598// CHECK1:       omp.inner.for.body:599// CHECK1-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4600// CHECK1-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP10]], 1601// CHECK1-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]602// CHECK1-NEXT:    store i32 [[ADD]], ptr [[I]], align 4603// CHECK1-NEXT:    [[TMP11:%.*]] = load i32, ptr [[I]], align 4604// CHECK1-NEXT:    [[TMP12:%.*]] = load i32, ptr [[T_VAR2]], align 4605// CHECK1-NEXT:    [[ADD4:%.*]] = add nsw i32 [[TMP12]], [[TMP11]]606// CHECK1-NEXT:    store i32 [[ADD4]], ptr [[T_VAR2]], align 4607// CHECK1-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]608// CHECK1:       omp.body.continue:609// CHECK1-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]610// CHECK1:       omp.inner.for.inc:611// CHECK1-NEXT:    [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4612// CHECK1-NEXT:    [[ADD5:%.*]] = add nsw i32 [[TMP13]], 1613// CHECK1-NEXT:    store i32 [[ADD5]], ptr [[DOTOMP_IV]], align 4614// CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND]]615// CHECK1:       omp.inner.for.end:616// CHECK1-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]617// CHECK1:       omp.loop.exit:618// CHECK1-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP4]])619// CHECK1-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i64 0, i64 0620// CHECK1-NEXT:    store ptr [[T_VAR2]], ptr [[TMP14]], align 8621// CHECK1-NEXT:    [[TMP15:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], i32 1, i64 8, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)622// CHECK1-NEXT:    switch i32 [[TMP15]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [623// CHECK1-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]624// CHECK1-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]625// CHECK1-NEXT:    ]626// CHECK1:       .omp.reduction.case1:627// CHECK1-NEXT:    [[TMP16:%.*]] = load i32, ptr [[TMP0]], align 4628// CHECK1-NEXT:    [[TMP17:%.*]] = load i32, ptr [[T_VAR2]], align 4629// CHECK1-NEXT:    [[ADD6:%.*]] = add nsw i32 [[TMP16]], [[TMP17]]630// CHECK1-NEXT:    store i32 [[ADD6]], ptr [[TMP0]], align 4631// CHECK1-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], ptr @.gomp_critical_user_.reduction.var)632// CHECK1-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]633// CHECK1:       .omp.reduction.case2:634// CHECK1-NEXT:    [[TMP18:%.*]] = load i32, ptr [[T_VAR2]], align 4635// CHECK1-NEXT:    [[TMP19:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP18]] monotonic, align 4636// CHECK1-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]637// CHECK1:       .omp.reduction.default:638// CHECK1-NEXT:    ret void639//640//641// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp_outlined.omp.reduction.reduction_func642// CHECK1-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR3]] {643// CHECK1-NEXT:  entry:644// CHECK1-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 8645// CHECK1-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 8646// CHECK1-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 8647// CHECK1-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 8648// CHECK1-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 8649// CHECK1-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 8650// CHECK1-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i64 0, i64 0651// CHECK1-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 8652// CHECK1-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i64 0, i64 0653// CHECK1-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 8654// CHECK1-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 4655// CHECK1-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 4656// CHECK1-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]657// CHECK1-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 4658// CHECK1-NEXT:    ret void659//660//661// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp.reduction.reduction_func662// CHECK1-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR3]] {663// CHECK1-NEXT:  entry:664// CHECK1-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 8665// CHECK1-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 8666// CHECK1-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 8667// CHECK1-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 8668// CHECK1-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 8669// CHECK1-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 8670// CHECK1-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i64 0, i64 0671// CHECK1-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 8672// CHECK1-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i64 0, i64 0673// CHECK1-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 8674// CHECK1-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 4675// CHECK1-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 4676// CHECK1-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]677// CHECK1-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 4678// CHECK1-NEXT:    ret void679//680//681// CHECK3-LABEL: define {{[^@]+}}@main682// CHECK3-SAME: () #[[ATTR0:[0-9]+]] {683// CHECK3-NEXT:  entry:684// CHECK3-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4685// CHECK3-NEXT:    [[SIVAR_CASTED:%.*]] = alloca i32, align 4686// CHECK3-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 4687// CHECK3-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 4688// CHECK3-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 4689// CHECK3-NEXT:    [[TMP:%.*]] = alloca i32, align 4690// CHECK3-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8691// CHECK3-NEXT:    store i32 0, ptr [[RETVAL]], align 4692// CHECK3-NEXT:    [[TMP0:%.*]] = load i32, ptr @_ZZ4mainE5sivar, align 4693// CHECK3-NEXT:    store i32 [[TMP0]], ptr [[SIVAR_CASTED]], align 4694// CHECK3-NEXT:    [[TMP1:%.*]] = load i32, ptr [[SIVAR_CASTED]], align 4695// CHECK3-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0696// CHECK3-NEXT:    store i32 [[TMP1]], ptr [[TMP2]], align 4697// CHECK3-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0698// CHECK3-NEXT:    store i32 [[TMP1]], ptr [[TMP3]], align 4699// CHECK3-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0700// CHECK3-NEXT:    store ptr null, ptr [[TMP4]], align 4701// CHECK3-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0702// CHECK3-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0703// CHECK3-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0704// CHECK3-NEXT:    store i32 3, ptr [[TMP7]], align 4705// CHECK3-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1706// CHECK3-NEXT:    store i32 1, ptr [[TMP8]], align 4707// CHECK3-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2708// CHECK3-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 4709// CHECK3-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3710// CHECK3-NEXT:    store ptr [[TMP6]], ptr [[TMP10]], align 4711// CHECK3-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4712// CHECK3-NEXT:    store ptr @.offload_sizes, ptr [[TMP11]], align 4713// CHECK3-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5714// CHECK3-NEXT:    store ptr @.offload_maptypes, ptr [[TMP12]], align 4715// CHECK3-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6716// CHECK3-NEXT:    store ptr null, ptr [[TMP13]], align 4717// CHECK3-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7718// CHECK3-NEXT:    store ptr null, ptr [[TMP14]], align 4719// CHECK3-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8720// CHECK3-NEXT:    store i64 2, ptr [[TMP15]], align 8721// CHECK3-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9722// CHECK3-NEXT:    store i64 0, ptr [[TMP16]], align 8723// CHECK3-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10724// CHECK3-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP17]], align 4725// CHECK3-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11726// CHECK3-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP18]], align 4727// CHECK3-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12728// CHECK3-NEXT:    store i32 0, ptr [[TMP19]], align 4729// CHECK3-NEXT:    [[TMP20:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB4:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.region_id, ptr [[KERNEL_ARGS]])730// CHECK3-NEXT:    [[TMP21:%.*]] = icmp ne i32 [[TMP20]], 0731// CHECK3-NEXT:    br i1 [[TMP21]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]732// CHECK3:       omp_offload.failed:733// CHECK3-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68(i32 [[TMP1]]) #[[ATTR2:[0-9]+]]734// CHECK3-NEXT:    br label [[OMP_OFFLOAD_CONT]]735// CHECK3:       omp_offload.cont:736// CHECK3-NEXT:    [[CALL:%.*]] = call noundef i32 @_Z5tmainIiET_v()737// CHECK3-NEXT:    ret i32 [[CALL]]738//739//740// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68741// CHECK3-SAME: (i32 noundef [[SIVAR:%.*]]) #[[ATTR1:[0-9]+]] {742// CHECK3-NEXT:  entry:743// CHECK3-NEXT:    [[SIVAR_ADDR:%.*]] = alloca i32, align 4744// CHECK3-NEXT:    store i32 [[SIVAR]], ptr [[SIVAR_ADDR]], align 4745// CHECK3-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB4]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined, ptr [[SIVAR_ADDR]])746// CHECK3-NEXT:    ret void747//748//749// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined750// CHECK3-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[SIVAR:%.*]]) #[[ATTR1]] {751// CHECK3-NEXT:  entry:752// CHECK3-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4753// CHECK3-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4754// CHECK3-NEXT:    [[SIVAR_ADDR:%.*]] = alloca ptr, align 4755// CHECK3-NEXT:    [[SIVAR1:%.*]] = alloca i32, align 4756// CHECK3-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4757// CHECK3-NEXT:    [[TMP:%.*]] = alloca i32, align 4758// CHECK3-NEXT:    [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4759// CHECK3-NEXT:    [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4760// CHECK3-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4761// CHECK3-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4762// CHECK3-NEXT:    [[I:%.*]] = alloca i32, align 4763// CHECK3-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 4764// CHECK3-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4765// CHECK3-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4766// CHECK3-NEXT:    store ptr [[SIVAR]], ptr [[SIVAR_ADDR]], align 4767// CHECK3-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[SIVAR_ADDR]], align 4768// CHECK3-NEXT:    store i32 0, ptr [[SIVAR1]], align 4769// CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_COMB_LB]], align 4770// CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_COMB_UB]], align 4771// CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4772// CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4773// CHECK3-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4774// CHECK3-NEXT:    [[TMP2:%.*]] = load i32, ptr [[TMP1]], align 4775// CHECK3-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP2]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)776// CHECK3-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4777// CHECK3-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP3]], 1778// CHECK3-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]779// CHECK3:       cond.true:780// CHECK3-NEXT:    br label [[COND_END:%.*]]781// CHECK3:       cond.false:782// CHECK3-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4783// CHECK3-NEXT:    br label [[COND_END]]784// CHECK3:       cond.end:785// CHECK3-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP4]], [[COND_FALSE]] ]786// CHECK3-NEXT:    store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4787// CHECK3-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4788// CHECK3-NEXT:    store i32 [[TMP5]], ptr [[DOTOMP_IV]], align 4789// CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]790// CHECK3:       omp.inner.for.cond:791// CHECK3-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4792// CHECK3-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4793// CHECK3-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP6]], [[TMP7]]794// CHECK3-NEXT:    br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]795// CHECK3:       omp.inner.for.body:796// CHECK3-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4797// CHECK3-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4798// CHECK3-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB4]], i32 3, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp_outlined, i32 [[TMP8]], i32 [[TMP9]], ptr [[SIVAR1]])799// CHECK3-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]800// CHECK3:       omp.inner.for.inc:801// CHECK3-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4802// CHECK3-NEXT:    [[TMP11:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4803// CHECK3-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP10]], [[TMP11]]804// CHECK3-NEXT:    store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4805// CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND]]806// CHECK3:       omp.inner.for.end:807// CHECK3-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]808// CHECK3:       omp.loop.exit:809// CHECK3-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP2]])810// CHECK3-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i32 0, i32 0811// CHECK3-NEXT:    store ptr [[SIVAR1]], ptr [[TMP12]], align 4812// CHECK3-NEXT:    [[TMP13:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3:[0-9]+]], i32 [[TMP2]], i32 1, i32 4, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)813// CHECK3-NEXT:    switch i32 [[TMP13]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [814// CHECK3-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]815// CHECK3-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]816// CHECK3-NEXT:    ]817// CHECK3:       .omp.reduction.case1:818// CHECK3-NEXT:    [[TMP14:%.*]] = load i32, ptr [[TMP0]], align 4819// CHECK3-NEXT:    [[TMP15:%.*]] = load i32, ptr [[SIVAR1]], align 4820// CHECK3-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP14]], [[TMP15]]821// CHECK3-NEXT:    store i32 [[ADD3]], ptr [[TMP0]], align 4822// CHECK3-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP2]], ptr @.gomp_critical_user_.reduction.var)823// CHECK3-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]824// CHECK3:       .omp.reduction.case2:825// CHECK3-NEXT:    [[TMP16:%.*]] = load i32, ptr [[SIVAR1]], align 4826// CHECK3-NEXT:    [[TMP17:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP16]] monotonic, align 4827// CHECK3-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]828// CHECK3:       .omp.reduction.default:829// CHECK3-NEXT:    ret void830//831//832// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp_outlined833// 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(4) [[SIVAR:%.*]]) #[[ATTR1]] {834// CHECK3-NEXT:  entry:835// CHECK3-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4836// CHECK3-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4837// CHECK3-NEXT:    [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i32, align 4838// CHECK3-NEXT:    [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i32, align 4839// CHECK3-NEXT:    [[SIVAR_ADDR:%.*]] = alloca ptr, align 4840// CHECK3-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4841// CHECK3-NEXT:    [[TMP:%.*]] = alloca i32, align 4842// CHECK3-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4843// CHECK3-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4844// CHECK3-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4845// CHECK3-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4846// CHECK3-NEXT:    [[SIVAR1:%.*]] = alloca i32, align 4847// CHECK3-NEXT:    [[I:%.*]] = alloca i32, align 4848// CHECK3-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 4849// CHECK3-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4850// CHECK3-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4851// CHECK3-NEXT:    store i32 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 4852// CHECK3-NEXT:    store i32 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 4853// CHECK3-NEXT:    store ptr [[SIVAR]], ptr [[SIVAR_ADDR]], align 4854// CHECK3-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[SIVAR_ADDR]], align 4855// CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4856// CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4857// CHECK3-NEXT:    [[TMP1:%.*]] = load i32, ptr [[DOTPREVIOUS_LB__ADDR]], align 4858// CHECK3-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTPREVIOUS_UB__ADDR]], align 4859// CHECK3-NEXT:    store i32 [[TMP1]], ptr [[DOTOMP_LB]], align 4860// CHECK3-NEXT:    store i32 [[TMP2]], ptr [[DOTOMP_UB]], align 4861// CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4862// CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4863// CHECK3-NEXT:    store i32 0, ptr [[SIVAR1]], align 4864// CHECK3-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4865// CHECK3-NEXT:    [[TMP4:%.*]] = load i32, ptr [[TMP3]], align 4866// CHECK3-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB2:[0-9]+]], i32 [[TMP4]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)867// CHECK3-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4868// CHECK3-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP5]], 1869// CHECK3-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]870// CHECK3:       cond.true:871// CHECK3-NEXT:    br label [[COND_END:%.*]]872// CHECK3:       cond.false:873// CHECK3-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4874// CHECK3-NEXT:    br label [[COND_END]]875// CHECK3:       cond.end:876// CHECK3-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP6]], [[COND_FALSE]] ]877// CHECK3-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4878// CHECK3-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4879// CHECK3-NEXT:    store i32 [[TMP7]], ptr [[DOTOMP_IV]], align 4880// CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]881// CHECK3:       omp.inner.for.cond:882// CHECK3-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4883// CHECK3-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4884// CHECK3-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP8]], [[TMP9]]885// CHECK3-NEXT:    br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]886// CHECK3:       omp.inner.for.body:887// CHECK3-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4888// CHECK3-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP10]], 1889// CHECK3-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]890// CHECK3-NEXT:    store i32 [[ADD]], ptr [[I]], align 4891// CHECK3-NEXT:    [[TMP11:%.*]] = load i32, ptr [[I]], align 4892// CHECK3-NEXT:    [[TMP12:%.*]] = load i32, ptr [[SIVAR1]], align 4893// CHECK3-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP12]], [[TMP11]]894// CHECK3-NEXT:    store i32 [[ADD3]], ptr [[SIVAR1]], align 4895// CHECK3-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]896// CHECK3:       omp.body.continue:897// CHECK3-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]898// CHECK3:       omp.inner.for.inc:899// CHECK3-NEXT:    [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4900// CHECK3-NEXT:    [[ADD4:%.*]] = add nsw i32 [[TMP13]], 1901// CHECK3-NEXT:    store i32 [[ADD4]], ptr [[DOTOMP_IV]], align 4902// CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND]]903// CHECK3:       omp.inner.for.end:904// CHECK3-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]905// CHECK3:       omp.loop.exit:906// CHECK3-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP4]])907// CHECK3-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i32 0, i32 0908// CHECK3-NEXT:    store ptr [[SIVAR1]], ptr [[TMP14]], align 4909// CHECK3-NEXT:    [[TMP15:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], i32 1, i32 4, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)910// CHECK3-NEXT:    switch i32 [[TMP15]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [911// CHECK3-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]912// CHECK3-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]913// CHECK3-NEXT:    ]914// CHECK3:       .omp.reduction.case1:915// CHECK3-NEXT:    [[TMP16:%.*]] = load i32, ptr [[TMP0]], align 4916// CHECK3-NEXT:    [[TMP17:%.*]] = load i32, ptr [[SIVAR1]], align 4917// CHECK3-NEXT:    [[ADD5:%.*]] = add nsw i32 [[TMP16]], [[TMP17]]918// CHECK3-NEXT:    store i32 [[ADD5]], ptr [[TMP0]], align 4919// CHECK3-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], ptr @.gomp_critical_user_.reduction.var)920// CHECK3-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]921// CHECK3:       .omp.reduction.case2:922// CHECK3-NEXT:    [[TMP18:%.*]] = load i32, ptr [[SIVAR1]], align 4923// CHECK3-NEXT:    [[TMP19:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP18]] monotonic, align 4924// CHECK3-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]925// CHECK3:       .omp.reduction.default:926// CHECK3-NEXT:    ret void927//928//929// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp_outlined.omp.reduction.reduction_func930// CHECK3-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR3:[0-9]+]] {931// CHECK3-NEXT:  entry:932// CHECK3-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 4933// CHECK3-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 4934// CHECK3-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 4935// CHECK3-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 4936// CHECK3-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 4937// CHECK3-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 4938// CHECK3-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i32 0, i32 0939// CHECK3-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 4940// CHECK3-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i32 0, i32 0941// CHECK3-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 4942// CHECK3-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 4943// CHECK3-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 4944// CHECK3-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]945// CHECK3-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 4946// CHECK3-NEXT:    ret void947//948//949// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined.omp.reduction.reduction_func950// CHECK3-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR3]] {951// CHECK3-NEXT:  entry:952// CHECK3-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 4953// CHECK3-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 4954// CHECK3-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 4955// CHECK3-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 4956// CHECK3-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 4957// CHECK3-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 4958// CHECK3-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i32 0, i32 0959// CHECK3-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 4960// CHECK3-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i32 0, i32 0961// CHECK3-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 4962// CHECK3-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 4963// CHECK3-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 4964// CHECK3-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]965// CHECK3-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 4966// CHECK3-NEXT:    ret void967//968//969// CHECK3-LABEL: define {{[^@]+}}@_Z5tmainIiET_v970// CHECK3-SAME: () #[[ATTR5:[0-9]+]] comdat {971// CHECK3-NEXT:  entry:972// CHECK3-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4973// CHECK3-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4974// CHECK3-NEXT:    [[T_VAR_CASTED:%.*]] = alloca i32, align 4975// CHECK3-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [1 x ptr], align 4976// CHECK3-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [1 x ptr], align 4977// CHECK3-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [1 x ptr], align 4978// CHECK3-NEXT:    [[TMP:%.*]] = alloca i32, align 4979// CHECK3-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8980// CHECK3-NEXT:    store i32 0, ptr [[T_VAR]], align 4981// CHECK3-NEXT:    call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[VEC]], ptr align 4 @__const._Z5tmainIiET_v.vec, i32 8, i1 false)982// CHECK3-NEXT:    [[TMP0:%.*]] = load i32, ptr [[T_VAR]], align 4983// CHECK3-NEXT:    store i32 [[TMP0]], ptr [[T_VAR_CASTED]], align 4984// CHECK3-NEXT:    [[TMP1:%.*]] = load i32, ptr [[T_VAR_CASTED]], align 4985// CHECK3-NEXT:    [[TMP2:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0986// CHECK3-NEXT:    store i32 [[TMP1]], ptr [[TMP2]], align 4987// CHECK3-NEXT:    [[TMP3:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0988// CHECK3-NEXT:    store i32 [[TMP1]], ptr [[TMP3]], align 4989// CHECK3-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0990// CHECK3-NEXT:    store ptr null, ptr [[TMP4]], align 4991// CHECK3-NEXT:    [[TMP5:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0992// CHECK3-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0993// CHECK3-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0994// CHECK3-NEXT:    store i32 3, ptr [[TMP7]], align 4995// CHECK3-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1996// CHECK3-NEXT:    store i32 1, ptr [[TMP8]], align 4997// CHECK3-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2998// CHECK3-NEXT:    store ptr [[TMP5]], ptr [[TMP9]], align 4999// CHECK3-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 31000// CHECK3-NEXT:    store ptr [[TMP6]], ptr [[TMP10]], align 41001// CHECK3-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 41002// CHECK3-NEXT:    store ptr @.offload_sizes.1, ptr [[TMP11]], align 41003// CHECK3-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 51004// CHECK3-NEXT:    store ptr @.offload_maptypes.2, ptr [[TMP12]], align 41005// CHECK3-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 61006// CHECK3-NEXT:    store ptr null, ptr [[TMP13]], align 41007// CHECK3-NEXT:    [[TMP14:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 71008// CHECK3-NEXT:    store ptr null, ptr [[TMP14]], align 41009// CHECK3-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 81010// CHECK3-NEXT:    store i64 2, ptr [[TMP15]], align 81011// CHECK3-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 91012// CHECK3-NEXT:    store i64 0, ptr [[TMP16]], align 81013// CHECK3-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 101014// CHECK3-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP17]], align 41015// CHECK3-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 111016// CHECK3-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP18]], align 41017// CHECK3-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 121018// CHECK3-NEXT:    store i32 0, ptr [[TMP19]], align 41019// CHECK3-NEXT:    [[TMP20:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB4]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.region_id, ptr [[KERNEL_ARGS]])1020// CHECK3-NEXT:    [[TMP21:%.*]] = icmp ne i32 [[TMP20]], 01021// CHECK3-NEXT:    br i1 [[TMP21]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]1022// CHECK3:       omp_offload.failed:1023// CHECK3-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32(i32 [[TMP1]]) #[[ATTR2]]1024// CHECK3-NEXT:    br label [[OMP_OFFLOAD_CONT]]1025// CHECK3:       omp_offload.cont:1026// CHECK3-NEXT:    ret i32 01027//1028//1029// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l321030// CHECK3-SAME: (i32 noundef [[T_VAR:%.*]]) #[[ATTR1]] {1031// CHECK3-NEXT:  entry:1032// CHECK3-NEXT:    [[T_VAR_ADDR:%.*]] = alloca i32, align 41033// CHECK3-NEXT:    store i32 [[T_VAR]], ptr [[T_VAR_ADDR]], align 41034// CHECK3-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB4]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined, ptr [[T_VAR_ADDR]])1035// CHECK3-NEXT:    ret void1036//1037//1038// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined1039// CHECK3-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[T_VAR:%.*]]) #[[ATTR1]] {1040// CHECK3-NEXT:  entry:1041// CHECK3-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 41042// CHECK3-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 41043// CHECK3-NEXT:    [[T_VAR_ADDR:%.*]] = alloca ptr, align 41044// CHECK3-NEXT:    [[T_VAR1:%.*]] = alloca i32, align 41045// CHECK3-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 41046// CHECK3-NEXT:    [[TMP:%.*]] = alloca i32, align 41047// CHECK3-NEXT:    [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 41048// CHECK3-NEXT:    [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 41049// CHECK3-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 41050// CHECK3-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 41051// CHECK3-NEXT:    [[I:%.*]] = alloca i32, align 41052// CHECK3-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 41053// CHECK3-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 41054// CHECK3-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 41055// CHECK3-NEXT:    store ptr [[T_VAR]], ptr [[T_VAR_ADDR]], align 41056// CHECK3-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[T_VAR_ADDR]], align 41057// CHECK3-NEXT:    store i32 0, ptr [[T_VAR1]], align 41058// CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_COMB_LB]], align 41059// CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_COMB_UB]], align 41060// CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 41061// CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 41062// CHECK3-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 41063// CHECK3-NEXT:    [[TMP2:%.*]] = load i32, ptr [[TMP1]], align 41064// CHECK3-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP2]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)1065// CHECK3-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41066// CHECK3-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP3]], 11067// CHECK3-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]1068// CHECK3:       cond.true:1069// CHECK3-NEXT:    br label [[COND_END:%.*]]1070// CHECK3:       cond.false:1071// CHECK3-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41072// CHECK3-NEXT:    br label [[COND_END]]1073// CHECK3:       cond.end:1074// CHECK3-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP4]], [[COND_FALSE]] ]1075// CHECK3-NEXT:    store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 41076// CHECK3-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 41077// CHECK3-NEXT:    store i32 [[TMP5]], ptr [[DOTOMP_IV]], align 41078// CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]1079// CHECK3:       omp.inner.for.cond:1080// CHECK3-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41081// CHECK3-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41082// CHECK3-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP6]], [[TMP7]]1083// CHECK3-NEXT:    br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]1084// CHECK3:       omp.inner.for.body:1085// CHECK3-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 41086// CHECK3-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41087// CHECK3-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB4]], i32 3, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp_outlined, i32 [[TMP8]], i32 [[TMP9]], ptr [[T_VAR1]])1088// CHECK3-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]1089// CHECK3:       omp.inner.for.inc:1090// CHECK3-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41091// CHECK3-NEXT:    [[TMP11:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 41092// CHECK3-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP10]], [[TMP11]]1093// CHECK3-NEXT:    store i32 [[ADD]], ptr [[DOTOMP_IV]], align 41094// CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND]]1095// CHECK3:       omp.inner.for.end:1096// CHECK3-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]1097// CHECK3:       omp.loop.exit:1098// CHECK3-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP2]])1099// CHECK3-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i32 0, i32 01100// CHECK3-NEXT:    store ptr [[T_VAR1]], ptr [[TMP12]], align 41101// CHECK3-NEXT:    [[TMP13:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP2]], i32 1, i32 4, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)1102// CHECK3-NEXT:    switch i32 [[TMP13]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [1103// CHECK3-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]1104// CHECK3-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]1105// CHECK3-NEXT:    ]1106// CHECK3:       .omp.reduction.case1:1107// CHECK3-NEXT:    [[TMP14:%.*]] = load i32, ptr [[TMP0]], align 41108// CHECK3-NEXT:    [[TMP15:%.*]] = load i32, ptr [[T_VAR1]], align 41109// CHECK3-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP14]], [[TMP15]]1110// CHECK3-NEXT:    store i32 [[ADD3]], ptr [[TMP0]], align 41111// CHECK3-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP2]], ptr @.gomp_critical_user_.reduction.var)1112// CHECK3-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]1113// CHECK3:       .omp.reduction.case2:1114// CHECK3-NEXT:    [[TMP16:%.*]] = load i32, ptr [[T_VAR1]], align 41115// CHECK3-NEXT:    [[TMP17:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP16]] monotonic, align 41116// CHECK3-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]1117// CHECK3:       .omp.reduction.default:1118// CHECK3-NEXT:    ret void1119//1120//1121// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp_outlined1122// 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(4) [[T_VAR:%.*]]) #[[ATTR1]] {1123// CHECK3-NEXT:  entry:1124// CHECK3-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 41125// CHECK3-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 41126// CHECK3-NEXT:    [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i32, align 41127// CHECK3-NEXT:    [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i32, align 41128// CHECK3-NEXT:    [[T_VAR_ADDR:%.*]] = alloca ptr, align 41129// CHECK3-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 41130// CHECK3-NEXT:    [[TMP:%.*]] = alloca i32, align 41131// CHECK3-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 41132// CHECK3-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 41133// CHECK3-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 41134// CHECK3-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 41135// CHECK3-NEXT:    [[T_VAR1:%.*]] = alloca i32, align 41136// CHECK3-NEXT:    [[I:%.*]] = alloca i32, align 41137// CHECK3-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 41138// CHECK3-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 41139// CHECK3-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 41140// CHECK3-NEXT:    store i32 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 41141// CHECK3-NEXT:    store i32 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 41142// CHECK3-NEXT:    store ptr [[T_VAR]], ptr [[T_VAR_ADDR]], align 41143// CHECK3-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[T_VAR_ADDR]], align 41144// CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 41145// CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 41146// CHECK3-NEXT:    [[TMP1:%.*]] = load i32, ptr [[DOTPREVIOUS_LB__ADDR]], align 41147// CHECK3-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTPREVIOUS_UB__ADDR]], align 41148// CHECK3-NEXT:    store i32 [[TMP1]], ptr [[DOTOMP_LB]], align 41149// CHECK3-NEXT:    store i32 [[TMP2]], ptr [[DOTOMP_UB]], align 41150// CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 41151// CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 41152// CHECK3-NEXT:    store i32 0, ptr [[T_VAR1]], align 41153// CHECK3-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 41154// CHECK3-NEXT:    [[TMP4:%.*]] = load i32, ptr [[TMP3]], align 41155// CHECK3-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB2]], i32 [[TMP4]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)1156// CHECK3-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41157// CHECK3-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP5]], 11158// CHECK3-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]1159// CHECK3:       cond.true:1160// CHECK3-NEXT:    br label [[COND_END:%.*]]1161// CHECK3:       cond.false:1162// CHECK3-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41163// CHECK3-NEXT:    br label [[COND_END]]1164// CHECK3:       cond.end:1165// CHECK3-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP6]], [[COND_FALSE]] ]1166// CHECK3-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 41167// CHECK3-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_LB]], align 41168// CHECK3-NEXT:    store i32 [[TMP7]], ptr [[DOTOMP_IV]], align 41169// CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]1170// CHECK3:       omp.inner.for.cond:1171// CHECK3-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41172// CHECK3-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41173// CHECK3-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP8]], [[TMP9]]1174// CHECK3-NEXT:    br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]1175// CHECK3:       omp.inner.for.body:1176// CHECK3-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41177// CHECK3-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP10]], 11178// CHECK3-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]1179// CHECK3-NEXT:    store i32 [[ADD]], ptr [[I]], align 41180// CHECK3-NEXT:    [[TMP11:%.*]] = load i32, ptr [[I]], align 41181// CHECK3-NEXT:    [[TMP12:%.*]] = load i32, ptr [[T_VAR1]], align 41182// CHECK3-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP12]], [[TMP11]]1183// CHECK3-NEXT:    store i32 [[ADD3]], ptr [[T_VAR1]], align 41184// CHECK3-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]1185// CHECK3:       omp.body.continue:1186// CHECK3-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]1187// CHECK3:       omp.inner.for.inc:1188// CHECK3-NEXT:    [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41189// CHECK3-NEXT:    [[ADD4:%.*]] = add nsw i32 [[TMP13]], 11190// CHECK3-NEXT:    store i32 [[ADD4]], ptr [[DOTOMP_IV]], align 41191// CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND]]1192// CHECK3:       omp.inner.for.end:1193// CHECK3-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]1194// CHECK3:       omp.loop.exit:1195// CHECK3-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP4]])1196// CHECK3-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i32 0, i32 01197// CHECK3-NEXT:    store ptr [[T_VAR1]], ptr [[TMP14]], align 41198// CHECK3-NEXT:    [[TMP15:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], i32 1, i32 4, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)1199// CHECK3-NEXT:    switch i32 [[TMP15]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [1200// CHECK3-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]1201// CHECK3-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]1202// CHECK3-NEXT:    ]1203// CHECK3:       .omp.reduction.case1:1204// CHECK3-NEXT:    [[TMP16:%.*]] = load i32, ptr [[TMP0]], align 41205// CHECK3-NEXT:    [[TMP17:%.*]] = load i32, ptr [[T_VAR1]], align 41206// CHECK3-NEXT:    [[ADD5:%.*]] = add nsw i32 [[TMP16]], [[TMP17]]1207// CHECK3-NEXT:    store i32 [[ADD5]], ptr [[TMP0]], align 41208// CHECK3-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], ptr @.gomp_critical_user_.reduction.var)1209// CHECK3-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]1210// CHECK3:       .omp.reduction.case2:1211// CHECK3-NEXT:    [[TMP18:%.*]] = load i32, ptr [[T_VAR1]], align 41212// CHECK3-NEXT:    [[TMP19:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP18]] monotonic, align 41213// CHECK3-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]1214// CHECK3:       .omp.reduction.default:1215// CHECK3-NEXT:    ret void1216//1217//1218// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp_outlined.omp.reduction.reduction_func1219// CHECK3-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR3]] {1220// CHECK3-NEXT:  entry:1221// CHECK3-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 41222// CHECK3-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 41223// CHECK3-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 41224// CHECK3-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 41225// CHECK3-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 41226// CHECK3-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 41227// CHECK3-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i32 0, i32 01228// CHECK3-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 41229// CHECK3-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i32 0, i32 01230// CHECK3-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 41231// CHECK3-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 41232// CHECK3-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 41233// CHECK3-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]1234// CHECK3-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 41235// CHECK3-NEXT:    ret void1236//1237//1238// CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l32.omp_outlined.omp.reduction.reduction_func1239// CHECK3-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR3]] {1240// CHECK3-NEXT:  entry:1241// CHECK3-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 41242// CHECK3-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 41243// CHECK3-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 41244// CHECK3-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 41245// CHECK3-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 41246// CHECK3-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 41247// CHECK3-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i32 0, i32 01248// CHECK3-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 41249// CHECK3-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i32 0, i32 01250// CHECK3-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 41251// CHECK3-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 41252// CHECK3-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 41253// CHECK3-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]1254// CHECK3-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 41255// CHECK3-NEXT:    ret void1256//1257//1258// CHECK9-LABEL: define {{[^@]+}}@main1259// CHECK9-SAME: () #[[ATTR0:[0-9]+]] {1260// CHECK9-NEXT:  entry:1261// CHECK9-NEXT:    [[RETVAL:%.*]] = alloca i32, align 41262// CHECK9-NEXT:    [[REF_TMP:%.*]] = alloca [[CLASS_ANON:%.*]], align 11263// CHECK9-NEXT:    store i32 0, ptr [[RETVAL]], align 41264// CHECK9-NEXT:    call void @"_ZZ4mainENK3$_0clEv"(ptr noundef nonnull align 1 dereferenceable(1) [[REF_TMP]])1265// CHECK9-NEXT:    ret i32 01266//1267//1268// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l451269// CHECK9-SAME: (i64 noundef [[SIVAR:%.*]]) #[[ATTR2:[0-9]+]] {1270// CHECK9-NEXT:  entry:1271// CHECK9-NEXT:    [[SIVAR_ADDR:%.*]] = alloca i64, align 81272// CHECK9-NEXT:    store i64 [[SIVAR]], ptr [[SIVAR_ADDR]], align 81273// CHECK9-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB4:[0-9]+]], i32 1, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l45.omp_outlined, ptr [[SIVAR_ADDR]])1274// CHECK9-NEXT:    ret void1275//1276//1277// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l45.omp_outlined1278// CHECK9-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[SIVAR:%.*]]) #[[ATTR2]] {1279// CHECK9-NEXT:  entry:1280// CHECK9-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 81281// CHECK9-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 81282// CHECK9-NEXT:    [[SIVAR_ADDR:%.*]] = alloca ptr, align 81283// CHECK9-NEXT:    [[SIVAR1:%.*]] = alloca i32, align 41284// CHECK9-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 41285// CHECK9-NEXT:    [[TMP:%.*]] = alloca i32, align 41286// CHECK9-NEXT:    [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 41287// CHECK9-NEXT:    [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 41288// CHECK9-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 41289// CHECK9-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 41290// CHECK9-NEXT:    [[I:%.*]] = alloca i32, align 41291// CHECK9-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 81292// CHECK9-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 81293// CHECK9-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 81294// CHECK9-NEXT:    store ptr [[SIVAR]], ptr [[SIVAR_ADDR]], align 81295// CHECK9-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[SIVAR_ADDR]], align 81296// CHECK9-NEXT:    store i32 0, ptr [[SIVAR1]], align 41297// CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_COMB_LB]], align 41298// CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_COMB_UB]], align 41299// CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 41300// CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 41301// CHECK9-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 81302// CHECK9-NEXT:    [[TMP2:%.*]] = load i32, ptr [[TMP1]], align 41303// CHECK9-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP2]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)1304// CHECK9-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41305// CHECK9-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP3]], 11306// CHECK9-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]1307// CHECK9:       cond.true:1308// CHECK9-NEXT:    br label [[COND_END:%.*]]1309// CHECK9:       cond.false:1310// CHECK9-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41311// CHECK9-NEXT:    br label [[COND_END]]1312// CHECK9:       cond.end:1313// CHECK9-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP4]], [[COND_FALSE]] ]1314// CHECK9-NEXT:    store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 41315// CHECK9-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 41316// CHECK9-NEXT:    store i32 [[TMP5]], ptr [[DOTOMP_IV]], align 41317// CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]1318// CHECK9:       omp.inner.for.cond:1319// CHECK9-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41320// CHECK9-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41321// CHECK9-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP6]], [[TMP7]]1322// CHECK9-NEXT:    br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]1323// CHECK9:       omp.inner.for.body:1324// CHECK9-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 41325// CHECK9-NEXT:    [[TMP9:%.*]] = zext i32 [[TMP8]] to i641326// CHECK9-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 41327// CHECK9-NEXT:    [[TMP11:%.*]] = zext i32 [[TMP10]] to i641328// CHECK9-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB4]], i32 3, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l45.omp_outlined.omp_outlined, i64 [[TMP9]], i64 [[TMP11]], ptr [[SIVAR1]])1329// CHECK9-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]1330// CHECK9:       omp.inner.for.inc:1331// CHECK9-NEXT:    [[TMP12:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41332// CHECK9-NEXT:    [[TMP13:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 41333// CHECK9-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP12]], [[TMP13]]1334// CHECK9-NEXT:    store i32 [[ADD]], ptr [[DOTOMP_IV]], align 41335// CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND]]1336// CHECK9:       omp.inner.for.end:1337// CHECK9-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]1338// CHECK9:       omp.loop.exit:1339// CHECK9-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP2]])1340// CHECK9-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i64 0, i64 01341// CHECK9-NEXT:    store ptr [[SIVAR1]], ptr [[TMP14]], align 81342// CHECK9-NEXT:    [[TMP15:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3:[0-9]+]], i32 [[TMP2]], i32 1, i64 8, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l45.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)1343// CHECK9-NEXT:    switch i32 [[TMP15]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [1344// CHECK9-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]1345// CHECK9-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]1346// CHECK9-NEXT:    ]1347// CHECK9:       .omp.reduction.case1:1348// CHECK9-NEXT:    [[TMP16:%.*]] = load i32, ptr [[TMP0]], align 41349// CHECK9-NEXT:    [[TMP17:%.*]] = load i32, ptr [[SIVAR1]], align 41350// CHECK9-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP16]], [[TMP17]]1351// CHECK9-NEXT:    store i32 [[ADD3]], ptr [[TMP0]], align 41352// CHECK9-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP2]], ptr @.gomp_critical_user_.reduction.var)1353// CHECK9-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]1354// CHECK9:       .omp.reduction.case2:1355// CHECK9-NEXT:    [[TMP18:%.*]] = load i32, ptr [[SIVAR1]], align 41356// CHECK9-NEXT:    [[TMP19:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP18]] monotonic, align 41357// CHECK9-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]1358// CHECK9:       .omp.reduction.default:1359// CHECK9-NEXT:    ret void1360//1361//1362// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l45.omp_outlined.omp_outlined1363// 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(4) [[SIVAR:%.*]]) #[[ATTR2]] {1364// CHECK9-NEXT:  entry:1365// CHECK9-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 81366// CHECK9-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 81367// CHECK9-NEXT:    [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 81368// CHECK9-NEXT:    [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 81369// CHECK9-NEXT:    [[SIVAR_ADDR:%.*]] = alloca ptr, align 81370// CHECK9-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 41371// CHECK9-NEXT:    [[TMP:%.*]] = alloca i32, align 41372// CHECK9-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 41373// CHECK9-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 41374// CHECK9-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 41375// CHECK9-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 41376// CHECK9-NEXT:    [[SIVAR2:%.*]] = alloca i32, align 41377// CHECK9-NEXT:    [[I:%.*]] = alloca i32, align 41378// CHECK9-NEXT:    [[REF_TMP:%.*]] = alloca [[CLASS_ANON_0:%.*]], align 81379// CHECK9-NEXT:    [[DOTOMP_REDUCTION_RED_LIST:%.*]] = alloca [1 x ptr], align 81380// CHECK9-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 81381// CHECK9-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 81382// CHECK9-NEXT:    store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 81383// CHECK9-NEXT:    store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 81384// CHECK9-NEXT:    store ptr [[SIVAR]], ptr [[SIVAR_ADDR]], align 81385// CHECK9-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[SIVAR_ADDR]], align 81386// CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 41387// CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 41388// CHECK9-NEXT:    [[TMP1:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 81389// CHECK9-NEXT:    [[CONV:%.*]] = trunc i64 [[TMP1]] to i321390// CHECK9-NEXT:    [[TMP2:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 81391// CHECK9-NEXT:    [[CONV1:%.*]] = trunc i64 [[TMP2]] to i321392// CHECK9-NEXT:    store i32 [[CONV]], ptr [[DOTOMP_LB]], align 41393// CHECK9-NEXT:    store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 41394// CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 41395// CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 41396// CHECK9-NEXT:    store i32 0, ptr [[SIVAR2]], align 41397// CHECK9-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 81398// CHECK9-NEXT:    [[TMP4:%.*]] = load i32, ptr [[TMP3]], align 41399// CHECK9-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB2:[0-9]+]], i32 [[TMP4]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)1400// CHECK9-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41401// CHECK9-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP5]], 11402// CHECK9-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]1403// CHECK9:       cond.true:1404// CHECK9-NEXT:    br label [[COND_END:%.*]]1405// CHECK9:       cond.false:1406// CHECK9-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41407// CHECK9-NEXT:    br label [[COND_END]]1408// CHECK9:       cond.end:1409// CHECK9-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP6]], [[COND_FALSE]] ]1410// CHECK9-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 41411// CHECK9-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_LB]], align 41412// CHECK9-NEXT:    store i32 [[TMP7]], ptr [[DOTOMP_IV]], align 41413// CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]1414// CHECK9:       omp.inner.for.cond:1415// CHECK9-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41416// CHECK9-NEXT:    [[TMP9:%.*]] = load i32, ptr [[DOTOMP_UB]], align 41417// CHECK9-NEXT:    [[CMP3:%.*]] = icmp sle i32 [[TMP8]], [[TMP9]]1418// CHECK9-NEXT:    br i1 [[CMP3]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]1419// CHECK9:       omp.inner.for.body:1420// CHECK9-NEXT:    [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41421// CHECK9-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP10]], 11422// CHECK9-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]1423// CHECK9-NEXT:    store i32 [[ADD]], ptr [[I]], align 41424// CHECK9-NEXT:    [[TMP11:%.*]] = load i32, ptr [[I]], align 41425// CHECK9-NEXT:    [[TMP12:%.*]] = load i32, ptr [[SIVAR2]], align 41426// CHECK9-NEXT:    [[ADD4:%.*]] = add nsw i32 [[TMP12]], [[TMP11]]1427// CHECK9-NEXT:    store i32 [[ADD4]], ptr [[SIVAR2]], align 41428// CHECK9-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 01429// CHECK9-NEXT:    store ptr [[SIVAR2]], ptr [[TMP13]], align 81430// CHECK9-NEXT:    call void @"_ZZZ4mainENK3$_0clEvENKUlvE_clEv"(ptr noundef nonnull align 8 dereferenceable(8) [[REF_TMP]])1431// CHECK9-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]1432// CHECK9:       omp.body.continue:1433// CHECK9-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]1434// CHECK9:       omp.inner.for.inc:1435// CHECK9-NEXT:    [[TMP14:%.*]] = load i32, ptr [[DOTOMP_IV]], align 41436// CHECK9-NEXT:    [[ADD5:%.*]] = add nsw i32 [[TMP14]], 11437// CHECK9-NEXT:    store i32 [[ADD5]], ptr [[DOTOMP_IV]], align 41438// CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND]]1439// CHECK9:       omp.inner.for.end:1440// CHECK9-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]1441// CHECK9:       omp.loop.exit:1442// CHECK9-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP4]])1443// CHECK9-NEXT:    [[TMP15:%.*]] = getelementptr inbounds [1 x ptr], ptr [[DOTOMP_REDUCTION_RED_LIST]], i64 0, i64 01444// CHECK9-NEXT:    store ptr [[SIVAR2]], ptr [[TMP15]], align 81445// CHECK9-NEXT:    [[TMP16:%.*]] = call i32 @__kmpc_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], i32 1, i64 8, ptr [[DOTOMP_REDUCTION_RED_LIST]], ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l45.omp_outlined.omp_outlined.omp.reduction.reduction_func, ptr @.gomp_critical_user_.reduction.var)1446// CHECK9-NEXT:    switch i32 [[TMP16]], label [[DOTOMP_REDUCTION_DEFAULT:%.*]] [1447// CHECK9-NEXT:      i32 1, label [[DOTOMP_REDUCTION_CASE1:%.*]]1448// CHECK9-NEXT:      i32 2, label [[DOTOMP_REDUCTION_CASE2:%.*]]1449// CHECK9-NEXT:    ]1450// CHECK9:       .omp.reduction.case1:1451// CHECK9-NEXT:    [[TMP17:%.*]] = load i32, ptr [[TMP0]], align 41452// CHECK9-NEXT:    [[TMP18:%.*]] = load i32, ptr [[SIVAR2]], align 41453// CHECK9-NEXT:    [[ADD6:%.*]] = add nsw i32 [[TMP17]], [[TMP18]]1454// CHECK9-NEXT:    store i32 [[ADD6]], ptr [[TMP0]], align 41455// CHECK9-NEXT:    call void @__kmpc_end_reduce_nowait(ptr @[[GLOB3]], i32 [[TMP4]], ptr @.gomp_critical_user_.reduction.var)1456// CHECK9-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]1457// CHECK9:       .omp.reduction.case2:1458// CHECK9-NEXT:    [[TMP19:%.*]] = load i32, ptr [[SIVAR2]], align 41459// CHECK9-NEXT:    [[TMP20:%.*]] = atomicrmw add ptr [[TMP0]], i32 [[TMP19]] monotonic, align 41460// CHECK9-NEXT:    br label [[DOTOMP_REDUCTION_DEFAULT]]1461// CHECK9:       .omp.reduction.default:1462// CHECK9-NEXT:    ret void1463//1464//1465// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l45.omp_outlined.omp_outlined.omp.reduction.reduction_func1466// CHECK9-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR4:[0-9]+]] {1467// CHECK9-NEXT:  entry:1468// CHECK9-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 81469// CHECK9-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 81470// CHECK9-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 81471// CHECK9-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 81472// CHECK9-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 81473// CHECK9-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 81474// CHECK9-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i64 0, i64 01475// CHECK9-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 81476// CHECK9-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i64 0, i64 01477// CHECK9-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 81478// CHECK9-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 41479// CHECK9-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 41480// CHECK9-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]1481// CHECK9-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 41482// CHECK9-NEXT:    ret void1483//1484//1485// CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l45.omp_outlined.omp.reduction.reduction_func1486// CHECK9-SAME: (ptr noundef [[TMP0:%.*]], ptr noundef [[TMP1:%.*]]) #[[ATTR4]] {1487// CHECK9-NEXT:  entry:1488// CHECK9-NEXT:    [[DOTADDR:%.*]] = alloca ptr, align 81489// CHECK9-NEXT:    [[DOTADDR1:%.*]] = alloca ptr, align 81490// CHECK9-NEXT:    store ptr [[TMP0]], ptr [[DOTADDR]], align 81491// CHECK9-NEXT:    store ptr [[TMP1]], ptr [[DOTADDR1]], align 81492// CHECK9-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[DOTADDR]], align 81493// CHECK9-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[DOTADDR1]], align 81494// CHECK9-NEXT:    [[TMP4:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP3]], i64 0, i64 01495// CHECK9-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP4]], align 81496// CHECK9-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [1 x ptr], ptr [[TMP2]], i64 0, i64 01497// CHECK9-NEXT:    [[TMP7:%.*]] = load ptr, ptr [[TMP6]], align 81498// CHECK9-NEXT:    [[TMP8:%.*]] = load i32, ptr [[TMP7]], align 41499// CHECK9-NEXT:    [[TMP9:%.*]] = load i32, ptr [[TMP5]], align 41500// CHECK9-NEXT:    [[ADD:%.*]] = add nsw i32 [[TMP8]], [[TMP9]]1501// CHECK9-NEXT:    store i32 [[ADD]], ptr [[TMP7]], align 41502// CHECK9-NEXT:    ret void1503//1504