589 lines · cpp
1// NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs --replace-value-regex "__omp_offloading_[0-9a-z]+_[0-9a-z]+" "reduction_size[.].+[.]" "pl_cond[.].+[.|,]" --prefix-filecheck-ir-name _2// add -fopenmp-targets3 4// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix=CHECK15// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s6// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix=CHECK17 8// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -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}}"9// RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s10// RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"11// expected-no-diagnostics12#ifndef HEADER13#define HEADER14 15typedef __INTPTR_TYPE__ intptr_t;16 17 18void foo();19 20struct S {21 intptr_t a, b, c;22 S(intptr_t a) : a(a) {}23 operator char() { return a; }24 ~S() {}25};26 27template <typename T>28T tmain() {29#pragma omp target30#pragma omp teams31#pragma omp distribute parallel for proc_bind(master)32 for(int i = 0; i < 1000; i++) {}33 return T();34}35 36int main() {37#pragma omp target38#pragma omp teams39#pragma omp distribute parallel for proc_bind(spread)40 for(int i = 0; i < 1000; i++) {}41#pragma omp target42#pragma omp teams43#pragma omp distribute parallel for proc_bind(close)44 for(int i = 0; i < 1000; i++) {}45 return tmain<int>();46}47 48 49 50 51 52 53 54 55#endif56// CHECK1-LABEL: define {{[^@]+}}@main57// CHECK1-SAME: () #[[ATTR0:[0-9]+]] {58// CHECK1-NEXT: entry:59// CHECK1-NEXT: [[RETVAL:%.*]] = alloca i32, align 460// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 461// CHECK1-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 862// CHECK1-NEXT: [[_TMP1:%.*]] = alloca i32, align 463// CHECK1-NEXT: [[KERNEL_ARGS2:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 864// CHECK1-NEXT: store i32 0, ptr [[RETVAL]], align 465// CHECK1-NEXT: [[TMP0:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 066// CHECK1-NEXT: store i32 3, ptr [[TMP0]], align 467// CHECK1-NEXT: [[TMP1:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 168// CHECK1-NEXT: store i32 0, ptr [[TMP1]], align 469// CHECK1-NEXT: [[TMP2:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 270// CHECK1-NEXT: store ptr null, ptr [[TMP2]], align 871// CHECK1-NEXT: [[TMP3:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 372// CHECK1-NEXT: store ptr null, ptr [[TMP3]], align 873// CHECK1-NEXT: [[TMP4:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 474// CHECK1-NEXT: store ptr null, ptr [[TMP4]], align 875// CHECK1-NEXT: [[TMP5:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 576// CHECK1-NEXT: store ptr null, ptr [[TMP5]], align 877// CHECK1-NEXT: [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 678// CHECK1-NEXT: store ptr null, ptr [[TMP6]], align 879// CHECK1-NEXT: [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 780// CHECK1-NEXT: store ptr null, ptr [[TMP7]], align 881// CHECK1-NEXT: [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 882// CHECK1-NEXT: store i64 1000, ptr [[TMP8]], align 883// CHECK1-NEXT: [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 984// CHECK1-NEXT: store i64 0, ptr [[TMP9]], align 885// CHECK1-NEXT: [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1086// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP10]], align 487// CHECK1-NEXT: [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1188// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP11]], align 489// CHECK1-NEXT: [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1290// CHECK1-NEXT: store i32 0, ptr [[TMP12]], align 491// CHECK1-NEXT: [[TMP13:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB3:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l37.region_id, ptr [[KERNEL_ARGS]])92// CHECK1-NEXT: [[TMP14:%.*]] = icmp ne i32 [[TMP13]], 093// CHECK1-NEXT: br i1 [[TMP14]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]94// CHECK1: omp_offload.failed:95// CHECK1-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l37() #[[ATTR2:[0-9]+]]96// CHECK1-NEXT: br label [[OMP_OFFLOAD_CONT]]97// CHECK1: omp_offload.cont:98// CHECK1-NEXT: [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 099// CHECK1-NEXT: store i32 3, ptr [[TMP15]], align 4100// CHECK1-NEXT: [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 1101// CHECK1-NEXT: store i32 0, ptr [[TMP16]], align 4102// CHECK1-NEXT: [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 2103// CHECK1-NEXT: store ptr null, ptr [[TMP17]], align 8104// CHECK1-NEXT: [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 3105// CHECK1-NEXT: store ptr null, ptr [[TMP18]], align 8106// CHECK1-NEXT: [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 4107// CHECK1-NEXT: store ptr null, ptr [[TMP19]], align 8108// CHECK1-NEXT: [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 5109// CHECK1-NEXT: store ptr null, ptr [[TMP20]], align 8110// CHECK1-NEXT: [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 6111// CHECK1-NEXT: store ptr null, ptr [[TMP21]], align 8112// CHECK1-NEXT: [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 7113// CHECK1-NEXT: store ptr null, ptr [[TMP22]], align 8114// CHECK1-NEXT: [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 8115// CHECK1-NEXT: store i64 1000, ptr [[TMP23]], align 8116// CHECK1-NEXT: [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 9117// CHECK1-NEXT: store i64 0, ptr [[TMP24]], align 8118// CHECK1-NEXT: [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 10119// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP25]], align 4120// CHECK1-NEXT: [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 11121// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP26]], align 4122// CHECK1-NEXT: [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS2]], i32 0, i32 12123// CHECK1-NEXT: store i32 0, ptr [[TMP27]], align 4124// CHECK1-NEXT: [[TMP28:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB3]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l41.region_id, ptr [[KERNEL_ARGS2]])125// CHECK1-NEXT: [[TMP29:%.*]] = icmp ne i32 [[TMP28]], 0126// CHECK1-NEXT: br i1 [[TMP29]], label [[OMP_OFFLOAD_FAILED3:%.*]], label [[OMP_OFFLOAD_CONT4:%.*]]127// CHECK1: omp_offload.failed3:128// CHECK1-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l41() #[[ATTR2]]129// CHECK1-NEXT: br label [[OMP_OFFLOAD_CONT4]]130// CHECK1: omp_offload.cont4:131// CHECK1-NEXT: [[CALL:%.*]] = call noundef signext i32 @_Z5tmainIiET_v()132// CHECK1-NEXT: ret i32 [[CALL]]133//134//135// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l37136// CHECK1-SAME: () #[[ATTR1:[0-9]+]] {137// CHECK1-NEXT: entry:138// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB3]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l37.omp_outlined)139// CHECK1-NEXT: ret void140//141//142// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l37.omp_outlined143// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR1]] {144// CHECK1-NEXT: entry:145// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8146// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8147// CHECK1-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4148// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4149// CHECK1-NEXT: [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4150// CHECK1-NEXT: [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4151// CHECK1-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4152// CHECK1-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4153// CHECK1-NEXT: [[I:%.*]] = alloca i32, align 4154// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8155// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8156// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_COMB_LB]], align 4157// CHECK1-NEXT: store i32 999, ptr [[DOTOMP_COMB_UB]], align 4158// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4159// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4160// CHECK1-NEXT: [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8161// CHECK1-NEXT: [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4162// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)163// CHECK1-NEXT: [[TMP2:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4164// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 999165// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]166// CHECK1: cond.true:167// CHECK1-NEXT: br label [[COND_END:%.*]]168// CHECK1: cond.false:169// CHECK1-NEXT: [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4170// CHECK1-NEXT: br label [[COND_END]]171// CHECK1: cond.end:172// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 999, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]173// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4174// CHECK1-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4175// CHECK1-NEXT: store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4176// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]177// CHECK1: omp.inner.for.cond:178// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4179// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4180// CHECK1-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]181// CHECK1-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]182// CHECK1: omp.inner.for.body:183// CHECK1-NEXT: call void @__kmpc_push_proc_bind(ptr @[[GLOB3]], i32 [[TMP1]], i32 4)184// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4185// CHECK1-NEXT: [[TMP8:%.*]] = zext i32 [[TMP7]] to i64186// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4187// CHECK1-NEXT: [[TMP10:%.*]] = zext i32 [[TMP9]] to i64188// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB3]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l37.omp_outlined.omp_outlined, i64 [[TMP8]], i64 [[TMP10]])189// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]190// CHECK1: omp.inner.for.inc:191// CHECK1-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4192// CHECK1-NEXT: [[TMP12:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4193// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP11]], [[TMP12]]194// CHECK1-NEXT: store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4195// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]196// CHECK1: omp.inner.for.end:197// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]198// CHECK1: omp.loop.exit:199// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP1]])200// CHECK1-NEXT: ret void201//202//203// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l37.omp_outlined.omp_outlined204// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], i64 noundef [[DOTPREVIOUS_LB_:%.*]], i64 noundef [[DOTPREVIOUS_UB_:%.*]]) #[[ATTR1]] {205// CHECK1-NEXT: entry:206// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8207// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8208// CHECK1-NEXT: [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 8209// CHECK1-NEXT: [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 8210// CHECK1-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4211// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4212// CHECK1-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 4213// CHECK1-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 4214// CHECK1-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4215// CHECK1-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4216// CHECK1-NEXT: [[I:%.*]] = alloca i32, align 4217// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8218// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8219// CHECK1-NEXT: store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 8220// CHECK1-NEXT: store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 8221// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_LB]], align 4222// CHECK1-NEXT: store i32 999, ptr [[DOTOMP_UB]], align 4223// CHECK1-NEXT: [[TMP0:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 8224// CHECK1-NEXT: [[CONV:%.*]] = trunc i64 [[TMP0]] to i32225// CHECK1-NEXT: [[TMP1:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 8226// CHECK1-NEXT: [[CONV1:%.*]] = trunc i64 [[TMP1]] to i32227// CHECK1-NEXT: store i32 [[CONV]], ptr [[DOTOMP_LB]], align 4228// CHECK1-NEXT: store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 4229// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4230// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4231// CHECK1-NEXT: [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8232// CHECK1-NEXT: [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 4233// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB2:[0-9]+]], i32 [[TMP3]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)234// CHECK1-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4235// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 999236// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]237// CHECK1: cond.true:238// CHECK1-NEXT: br label [[COND_END:%.*]]239// CHECK1: cond.false:240// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4241// CHECK1-NEXT: br label [[COND_END]]242// CHECK1: cond.end:243// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 999, [[COND_TRUE]] ], [ [[TMP5]], [[COND_FALSE]] ]244// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_UB]], align 4245// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4246// CHECK1-NEXT: store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 4247// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]248// CHECK1: omp.inner.for.cond:249// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4250// CHECK1-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4251// CHECK1-NEXT: [[CMP2:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]252// CHECK1-NEXT: br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]253// CHECK1: omp.inner.for.body:254// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4255// CHECK1-NEXT: [[MUL:%.*]] = mul nsw i32 [[TMP9]], 1256// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 0, [[MUL]]257// CHECK1-NEXT: store i32 [[ADD]], ptr [[I]], align 4258// CHECK1-NEXT: br label [[OMP_BODY_CONTINUE:%.*]]259// CHECK1: omp.body.continue:260// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]261// CHECK1: omp.inner.for.inc:262// CHECK1-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4263// CHECK1-NEXT: [[ADD3:%.*]] = add nsw i32 [[TMP10]], 1264// CHECK1-NEXT: store i32 [[ADD3]], ptr [[DOTOMP_IV]], align 4265// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]266// CHECK1: omp.inner.for.end:267// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]268// CHECK1: omp.loop.exit:269// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP3]])270// CHECK1-NEXT: ret void271//272//273// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l41274// CHECK1-SAME: () #[[ATTR1]] {275// CHECK1-NEXT: entry:276// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB3]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l41.omp_outlined)277// CHECK1-NEXT: ret void278//279//280// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l41.omp_outlined281// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR1]] {282// CHECK1-NEXT: entry:283// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8284// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8285// CHECK1-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4286// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4287// CHECK1-NEXT: [[DOTOMP_COMB_LB:%.*]] = alloca i32, align 4288// CHECK1-NEXT: [[DOTOMP_COMB_UB:%.*]] = alloca i32, align 4289// CHECK1-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4290// CHECK1-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4291// CHECK1-NEXT: [[I:%.*]] = alloca i32, align 4292// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8293// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8294// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_COMB_LB]], align 4295// CHECK1-NEXT: store i32 999, ptr [[DOTOMP_COMB_UB]], align 4296// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4297// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4298// CHECK1-NEXT: [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8299// CHECK1-NEXT: [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4300// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)301// CHECK1-NEXT: [[TMP2:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4302// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 999303// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]304// CHECK1: cond.true:305// CHECK1-NEXT: br label [[COND_END:%.*]]306// CHECK1: cond.false:307// CHECK1-NEXT: [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4308// CHECK1-NEXT: br label [[COND_END]]309// CHECK1: cond.end:310// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 999, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]311// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4312// CHECK1-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4313// CHECK1-NEXT: store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4314// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]315// CHECK1: omp.inner.for.cond:316// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4317// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4318// CHECK1-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]319// CHECK1-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]320// CHECK1: omp.inner.for.body:321// CHECK1-NEXT: call void @__kmpc_push_proc_bind(ptr @[[GLOB3]], i32 [[TMP1]], i32 3)322// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4323// CHECK1-NEXT: [[TMP8:%.*]] = zext i32 [[TMP7]] to i64324// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4325// CHECK1-NEXT: [[TMP10:%.*]] = zext i32 [[TMP9]] to i64326// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB3]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l41.omp_outlined.omp_outlined, i64 [[TMP8]], i64 [[TMP10]])327// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]328// CHECK1: omp.inner.for.inc:329// CHECK1-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4330// CHECK1-NEXT: [[TMP12:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4331// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP11]], [[TMP12]]332// CHECK1-NEXT: store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4333// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]334// CHECK1: omp.inner.for.end:335// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]336// CHECK1: omp.loop.exit:337// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP1]])338// CHECK1-NEXT: ret void339//340//341// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l41.omp_outlined.omp_outlined342// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], i64 noundef [[DOTPREVIOUS_LB_:%.*]], i64 noundef [[DOTPREVIOUS_UB_:%.*]]) #[[ATTR1]] {343// CHECK1-NEXT: entry:344// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8345// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8346// CHECK1-NEXT: [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 8347// CHECK1-NEXT: [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 8348// CHECK1-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4349// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4350// CHECK1-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 4351// CHECK1-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 4352// CHECK1-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4353// CHECK1-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4354// CHECK1-NEXT: [[I:%.*]] = alloca i32, align 4355// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8356// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8357// CHECK1-NEXT: store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 8358// CHECK1-NEXT: store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 8359// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_LB]], align 4360// CHECK1-NEXT: store i32 999, ptr [[DOTOMP_UB]], align 4361// CHECK1-NEXT: [[TMP0:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 8362// CHECK1-NEXT: [[CONV:%.*]] = trunc i64 [[TMP0]] to i32363// CHECK1-NEXT: [[TMP1:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 8364// CHECK1-NEXT: [[CONV1:%.*]] = trunc i64 [[TMP1]] to i32365// CHECK1-NEXT: store i32 [[CONV]], ptr [[DOTOMP_LB]], align 4366// CHECK1-NEXT: store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 4367// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4368// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4369// CHECK1-NEXT: [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8370// CHECK1-NEXT: [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 4371// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB2]], i32 [[TMP3]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)372// CHECK1-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4373// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 999374// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]375// CHECK1: cond.true:376// CHECK1-NEXT: br label [[COND_END:%.*]]377// CHECK1: cond.false:378// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4379// CHECK1-NEXT: br label [[COND_END]]380// CHECK1: cond.end:381// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 999, [[COND_TRUE]] ], [ [[TMP5]], [[COND_FALSE]] ]382// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_UB]], align 4383// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4384// CHECK1-NEXT: store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 4385// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]386// CHECK1: omp.inner.for.cond:387// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4388// CHECK1-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4389// CHECK1-NEXT: [[CMP2:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]390// CHECK1-NEXT: br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]391// CHECK1: omp.inner.for.body:392// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4393// CHECK1-NEXT: [[MUL:%.*]] = mul nsw i32 [[TMP9]], 1394// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 0, [[MUL]]395// CHECK1-NEXT: store i32 [[ADD]], ptr [[I]], align 4396// CHECK1-NEXT: br label [[OMP_BODY_CONTINUE:%.*]]397// CHECK1: omp.body.continue:398// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]399// CHECK1: omp.inner.for.inc:400// CHECK1-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4401// CHECK1-NEXT: [[ADD3:%.*]] = add nsw i32 [[TMP10]], 1402// CHECK1-NEXT: store i32 [[ADD3]], ptr [[DOTOMP_IV]], align 4403// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]404// CHECK1: omp.inner.for.end:405// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]406// CHECK1: omp.loop.exit:407// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP3]])408// CHECK1-NEXT: ret void409//410//411// CHECK1-LABEL: define {{[^@]+}}@_Z5tmainIiET_v412// CHECK1-SAME: () #[[ATTR3:[0-9]+]] comdat {413// CHECK1-NEXT: entry:414// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4415// CHECK1-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8416// CHECK1-NEXT: [[TMP0:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0417// CHECK1-NEXT: store i32 3, ptr [[TMP0]], align 4418// CHECK1-NEXT: [[TMP1:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1419// CHECK1-NEXT: store i32 0, ptr [[TMP1]], align 4420// CHECK1-NEXT: [[TMP2:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2421// CHECK1-NEXT: store ptr null, ptr [[TMP2]], align 8422// CHECK1-NEXT: [[TMP3:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3423// CHECK1-NEXT: store ptr null, ptr [[TMP3]], align 8424// CHECK1-NEXT: [[TMP4:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4425// CHECK1-NEXT: store ptr null, ptr [[TMP4]], align 8426// CHECK1-NEXT: [[TMP5:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5427// CHECK1-NEXT: store ptr null, ptr [[TMP5]], align 8428// CHECK1-NEXT: [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6429// CHECK1-NEXT: store ptr null, ptr [[TMP6]], align 8430// CHECK1-NEXT: [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7431// CHECK1-NEXT: store ptr null, ptr [[TMP7]], align 8432// CHECK1-NEXT: [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8433// CHECK1-NEXT: store i64 1000, ptr [[TMP8]], align 8434// CHECK1-NEXT: [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9435// CHECK1-NEXT: store i64 0, ptr [[TMP9]], align 8436// CHECK1-NEXT: [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10437// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP10]], align 4438// CHECK1-NEXT: [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11439// CHECK1-NEXT: store [3 x i32] zeroinitializer, ptr [[TMP11]], align 4440// CHECK1-NEXT: [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12441// CHECK1-NEXT: store i32 0, ptr [[TMP12]], align 4442// CHECK1-NEXT: [[TMP13:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB3]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l29.region_id, ptr [[KERNEL_ARGS]])443// CHECK1-NEXT: [[TMP14:%.*]] = icmp ne i32 [[TMP13]], 0444// CHECK1-NEXT: br i1 [[TMP14]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]445// CHECK1: omp_offload.failed:446// CHECK1-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l29() #[[ATTR2]]447// CHECK1-NEXT: br label [[OMP_OFFLOAD_CONT]]448// CHECK1: omp_offload.cont:449// CHECK1-NEXT: ret i32 0450//451//452// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l29453// CHECK1-SAME: () #[[ATTR1]] {454// CHECK1-NEXT: entry:455// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB3]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l29.omp_outlined)456// CHECK1-NEXT: ret void457//458//459// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l29.omp_outlined460// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR1]] {461// CHECK1-NEXT: entry:462// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8463// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8464// 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: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8472// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8473// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_COMB_LB]], align 4474// CHECK1-NEXT: store i32 999, ptr [[DOTOMP_COMB_UB]], align 4475// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4476// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4477// CHECK1-NEXT: [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8478// CHECK1-NEXT: [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4479// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_COMB_LB]], ptr [[DOTOMP_COMB_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)480// CHECK1-NEXT: [[TMP2:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4481// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 999482// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]483// CHECK1: cond.true:484// CHECK1-NEXT: br label [[COND_END:%.*]]485// CHECK1: cond.false:486// CHECK1-NEXT: [[TMP3:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4487// CHECK1-NEXT: br label [[COND_END]]488// CHECK1: cond.end:489// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 999, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]490// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_COMB_UB]], align 4491// CHECK1-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4492// CHECK1-NEXT: store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4493// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]494// CHECK1: omp.inner.for.cond:495// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4496// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4497// CHECK1-NEXT: [[CMP1:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]498// CHECK1-NEXT: br i1 [[CMP1]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]499// CHECK1: omp.inner.for.body:500// CHECK1-NEXT: call void @__kmpc_push_proc_bind(ptr @[[GLOB3]], i32 [[TMP1]], i32 2)501// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_COMB_LB]], align 4502// CHECK1-NEXT: [[TMP8:%.*]] = zext i32 [[TMP7]] to i64503// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_COMB_UB]], align 4504// CHECK1-NEXT: [[TMP10:%.*]] = zext i32 [[TMP9]] to i64505// CHECK1-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @[[GLOB3]], i32 2, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l29.omp_outlined.omp_outlined, i64 [[TMP8]], i64 [[TMP10]])506// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]507// CHECK1: omp.inner.for.inc:508// CHECK1-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4509// CHECK1-NEXT: [[TMP12:%.*]] = load i32, ptr [[DOTOMP_STRIDE]], align 4510// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 [[TMP11]], [[TMP12]]511// CHECK1-NEXT: store i32 [[ADD]], ptr [[DOTOMP_IV]], align 4512// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]513// CHECK1: omp.inner.for.end:514// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]515// CHECK1: omp.loop.exit:516// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP1]])517// CHECK1-NEXT: ret void518//519//520// CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l29.omp_outlined.omp_outlined521// CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]], i64 noundef [[DOTPREVIOUS_LB_:%.*]], i64 noundef [[DOTPREVIOUS_UB_:%.*]]) #[[ATTR1]] {522// CHECK1-NEXT: entry:523// CHECK1-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8524// CHECK1-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8525// CHECK1-NEXT: [[DOTPREVIOUS_LB__ADDR:%.*]] = alloca i64, align 8526// CHECK1-NEXT: [[DOTPREVIOUS_UB__ADDR:%.*]] = alloca i64, align 8527// CHECK1-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4528// CHECK1-NEXT: [[TMP:%.*]] = alloca i32, align 4529// CHECK1-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 4530// CHECK1-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 4531// CHECK1-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4532// CHECK1-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4533// CHECK1-NEXT: [[I:%.*]] = alloca i32, align 4534// CHECK1-NEXT: store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8535// CHECK1-NEXT: store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8536// CHECK1-NEXT: store i64 [[DOTPREVIOUS_LB_]], ptr [[DOTPREVIOUS_LB__ADDR]], align 8537// CHECK1-NEXT: store i64 [[DOTPREVIOUS_UB_]], ptr [[DOTPREVIOUS_UB__ADDR]], align 8538// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_LB]], align 4539// CHECK1-NEXT: store i32 999, ptr [[DOTOMP_UB]], align 4540// CHECK1-NEXT: [[TMP0:%.*]] = load i64, ptr [[DOTPREVIOUS_LB__ADDR]], align 8541// CHECK1-NEXT: [[CONV:%.*]] = trunc i64 [[TMP0]] to i32542// CHECK1-NEXT: [[TMP1:%.*]] = load i64, ptr [[DOTPREVIOUS_UB__ADDR]], align 8543// CHECK1-NEXT: [[CONV1:%.*]] = trunc i64 [[TMP1]] to i32544// CHECK1-NEXT: store i32 [[CONV]], ptr [[DOTOMP_LB]], align 4545// CHECK1-NEXT: store i32 [[CONV1]], ptr [[DOTOMP_UB]], align 4546// CHECK1-NEXT: store i32 1, ptr [[DOTOMP_STRIDE]], align 4547// CHECK1-NEXT: store i32 0, ptr [[DOTOMP_IS_LAST]], align 4548// CHECK1-NEXT: [[TMP2:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8549// CHECK1-NEXT: [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 4550// CHECK1-NEXT: call void @__kmpc_for_static_init_4(ptr @[[GLOB2]], i32 [[TMP3]], i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)551// CHECK1-NEXT: [[TMP4:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4552// CHECK1-NEXT: [[CMP:%.*]] = icmp sgt i32 [[TMP4]], 999553// CHECK1-NEXT: br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]554// CHECK1: cond.true:555// CHECK1-NEXT: br label [[COND_END:%.*]]556// CHECK1: cond.false:557// CHECK1-NEXT: [[TMP5:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4558// CHECK1-NEXT: br label [[COND_END]]559// CHECK1: cond.end:560// CHECK1-NEXT: [[COND:%.*]] = phi i32 [ 999, [[COND_TRUE]] ], [ [[TMP5]], [[COND_FALSE]] ]561// CHECK1-NEXT: store i32 [[COND]], ptr [[DOTOMP_UB]], align 4562// CHECK1-NEXT: [[TMP6:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4563// CHECK1-NEXT: store i32 [[TMP6]], ptr [[DOTOMP_IV]], align 4564// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND:%.*]]565// CHECK1: omp.inner.for.cond:566// CHECK1-NEXT: [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4567// CHECK1-NEXT: [[TMP8:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4568// CHECK1-NEXT: [[CMP2:%.*]] = icmp sle i32 [[TMP7]], [[TMP8]]569// CHECK1-NEXT: br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]570// CHECK1: omp.inner.for.body:571// CHECK1-NEXT: [[TMP9:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4572// CHECK1-NEXT: [[MUL:%.*]] = mul nsw i32 [[TMP9]], 1573// CHECK1-NEXT: [[ADD:%.*]] = add nsw i32 0, [[MUL]]574// CHECK1-NEXT: store i32 [[ADD]], ptr [[I]], align 4575// CHECK1-NEXT: br label [[OMP_BODY_CONTINUE:%.*]]576// CHECK1: omp.body.continue:577// CHECK1-NEXT: br label [[OMP_INNER_FOR_INC:%.*]]578// CHECK1: omp.inner.for.inc:579// CHECK1-NEXT: [[TMP10:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4580// CHECK1-NEXT: [[ADD3:%.*]] = add nsw i32 [[TMP10]], 1581// CHECK1-NEXT: store i32 [[ADD3]], ptr [[DOTOMP_IV]], align 4582// CHECK1-NEXT: br label [[OMP_INNER_FOR_COND]]583// CHECK1: omp.inner.for.end:584// CHECK1-NEXT: br label [[OMP_LOOP_EXIT:%.*]]585// CHECK1: omp.loop.exit:586// CHECK1-NEXT: call void @__kmpc_for_static_fini(ptr @[[GLOB2]], i32 [[TMP3]])587// CHECK1-NEXT: ret void588//589