clang/test/OpenMP/distribute_private_codegen.cpp

   1 // NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs --replace-value-regex "__omp_offloading_[0-9a-z]+_[0-9a-z]+" "reduction_size[.].+[.]" "pl_cond[.].+[.|,]" --prefix-filecheck-ir-name _
   2 // RUN: %clang_cc1 -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=CHECK1
   3 // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
   4 // RUN: %clang_cc1 -DLAMBDA -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=CHECK1
   5 // RUN: %clang_cc1 -DLAMBDA -verify -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix=CHECK3
   6 // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s
   7 // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -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=CHECK3
   8
   9 // 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}}"
  10 // 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 %s
  11 // RUN: %clang_cc1 -DLAMBDA -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 -DLAMBDA -verify -fopenmp-simd -x c++ -std=c++11 -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 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s
  14 // RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"
  15
  16 // RUN: %clang_cc1  -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=CHECK9
  17 // RUN: %clang_cc1  -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
  18 // RUN: %clang_cc1  -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=CHECK9
  19 // RUN: %clang_cc1  -verify -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix=CHECK11
  20 // RUN: %clang_cc1  -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s
  21 // RUN: %clang_cc1  -fopenmp -x c++ -std=c++11 -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=CHECK11
  22
  23 // RUN: %clang_cc1  -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}}"
  24 // RUN: %clang_cc1  -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
  25 // RUN: %clang_cc1  -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}}"
  26 // RUN: %clang_cc1  -verify -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --implicit-check-not="{{__kmpc|__tgt}}"
  27 // RUN: %clang_cc1  -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s
  28 // RUN: %clang_cc1  -fopenmp-simd -x c++ -std=c++11 -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}}"
  29 // expected-no-diagnostics
  30 #ifndef HEADER
  31 #define HEADER
  32
  33 template <class T>
  34 struct S {
  35   T f;
  36   S(T a) : f(a) {}
  37   S() : f() {}
  38   operator T() { return T(); }
  39   ~S() {}
  40 };
  41
  42 template <typename T>
  43 T tmain() {
  44   S<T> test;
  45   T t_var = T();
  46   T vec[] = {1, 2};
  47   S<T> s_arr[] = {1, 2};
  48   S<T> &var = test;
  49   #pragma omp target
  50   #pragma omp teams
  51 #pragma omp distribute private(t_var, vec, s_arr, s_arr, var, var)
  52   for (int i = 0; i < 2; ++i) {
  53     vec[i] = t_var;
  54     s_arr[i] = var;
  55   }
  56   return T();
  57 }
  58
  59 int main() {
  60   static int svar;
  61   volatile double g;
  62   volatile double &g1 = g;
  63
  64   #ifdef LAMBDA
  65   [&]() {
  66     static float sfvar;
  67
  68     #pragma omp target
  69     #pragma omp teams
  70 #pragma omp distribute private(g, g1, svar, sfvar)
  71     for (int i = 0; i < 2; ++i) {
  72       g = 1;
  73       g1 = 1;
  74       svar = 3;
  75       sfvar = 4.0;
  76       [&]() {
  77         g = 2;
  78         g1 = 2;
  79         svar = 4;
  80         sfvar = 8.0;
  81
  82       }();
  83     }
  84   }();
  85   return 0;
  86   #else
  87   S<float> test;
  88   int t_var = 0;
  89   int vec[] = {1, 2};
  90   S<float> s_arr[] = {1, 2};
  91   S<float> &var = test;
  92
  93   #pragma omp target
  94   #pragma omp teams
  95 #pragma omp distribute private(t_var, vec, s_arr, s_arr, var, var, svar)
  96   for (int i = 0; i < 2; ++i) {
  97     vec[i] = t_var;
  98     s_arr[i] = var;
  99   }
 100   int i;
 101
 102   #pragma omp target
 103   #pragma omp teams
 104 #pragma omp distribute private(i)
 105   for (i = 0; i < 2; ++i) {
 106     ;
 107   }
 108   return tmain<int>();
 109   #endif
 110 }
 111
 112
 113
 114
 115
 116 #endif
 117 // CHECK1-LABEL: define {{[^@]+}}@main
 118 // CHECK1-SAME: () #[[ATTR0:[0-9]+]] {
 119 // CHECK1-NEXT:  entry:
 120 // CHECK1-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4
 121 // CHECK1-NEXT:    [[G:%.*]] = alloca double, align 8
 122 // CHECK1-NEXT:    [[G1:%.*]] = alloca ptr, align 8
 123 // CHECK1-NEXT:    [[REF_TMP:%.*]] = alloca [[CLASS_ANON:%.*]], align 1
 124 // CHECK1-NEXT:    store i32 0, ptr [[RETVAL]], align 4
 125 // CHECK1-NEXT:    store ptr [[G]], ptr [[G1]], align 8
 126 // CHECK1-NEXT:    call void @"_ZZ4mainENK3$_0clEv"(ptr noundef nonnull align 1 dereferenceable(1) [[REF_TMP]])
 127 // CHECK1-NEXT:    ret i32 0
 128 //
 129 //
 130 // CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68
 131 // CHECK1-SAME: () #[[ATTR2:[0-9]+]] {
 132 // CHECK1-NEXT:  entry:
 133 // CHECK1-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB2:[0-9]+]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined)
 134 // CHECK1-NEXT:    ret void
 135 //
 136 //
 137 // CHECK1-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined
 138 // CHECK1-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR2]] {
 139 // CHECK1-NEXT:  entry:
 140 // CHECK1-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8
 141 // CHECK1-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8
 142 // CHECK1-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4
 143 // CHECK1-NEXT:    [[TMP:%.*]] = alloca i32, align 4
 144 // CHECK1-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 8
 145 // CHECK1-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4
 146 // CHECK1-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4
 147 // CHECK1-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
 148 // CHECK1-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
 149 // CHECK1-NEXT:    [[G:%.*]] = alloca double, align 8
 150 // CHECK1-NEXT:    [[G1:%.*]] = alloca double, align 8
 151 // CHECK1-NEXT:    [[_TMP2:%.*]] = alloca ptr, align 8
 152 // CHECK1-NEXT:    [[SVAR:%.*]] = alloca i32, align 4
 153 // CHECK1-NEXT:    [[SFVAR:%.*]] = alloca float, align 4
 154 // CHECK1-NEXT:    [[I:%.*]] = alloca i32, align 4
 155 // CHECK1-NEXT:    [[REF_TMP:%.*]] = alloca [[CLASS_ANON_0:%.*]], align 8
 156 // CHECK1-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8
 157 // CHECK1-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8
 158 // CHECK1-NEXT:    store ptr undef, ptr [[_TMP1]], align 8
 159 // CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4
 160 // CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4
 161 // CHECK1-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4
 162 // CHECK1-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4
 163 // CHECK1-NEXT:    store ptr [[G1]], ptr [[_TMP2]], align 8
 164 // CHECK1-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8
 165 // CHECK1-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4
 166 // CHECK1-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
 167 // CHECK1-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 168 // CHECK1-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 1
 169 // CHECK1-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]
 170 // CHECK1:       cond.true:
 171 // CHECK1-NEXT:    br label [[COND_END:%.*]]
 172 // CHECK1:       cond.false:
 173 // CHECK1-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 174 // CHECK1-NEXT:    br label [[COND_END]]
 175 // CHECK1:       cond.end:
 176 // CHECK1-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]
 177 // CHECK1-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4
 178 // CHECK1-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4
 179 // CHECK1-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4
 180 // CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]
 181 // CHECK1:       omp.inner.for.cond:
 182 // CHECK1-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 183 // CHECK1-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 184 // CHECK1-NEXT:    [[CMP3:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]
 185 // CHECK1-NEXT:    br i1 [[CMP3]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]
 186 // CHECK1:       omp.inner.for.body:
 187 // CHECK1-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 188 // CHECK1-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP7]], 1
 189 // CHECK1-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]
 190 // CHECK1-NEXT:    store i32 [[ADD]], ptr [[I]], align 4
 191 // CHECK1-NEXT:    store double 1.000000e+00, ptr [[G]], align 8
 192 // CHECK1-NEXT:    [[TMP8:%.*]] = load ptr, ptr [[_TMP2]], align 8
 193 // CHECK1-NEXT:    store volatile double 1.000000e+00, ptr [[TMP8]], align 8
 194 // CHECK1-NEXT:    store i32 3, ptr [[SVAR]], align 4
 195 // CHECK1-NEXT:    store float 4.000000e+00, ptr [[SFVAR]], align 4
 196 // CHECK1-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 0
 197 // CHECK1-NEXT:    store ptr [[G]], ptr [[TMP9]], align 8
 198 // CHECK1-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 1
 199 // CHECK1-NEXT:    [[TMP11:%.*]] = load ptr, ptr [[_TMP2]], align 8
 200 // CHECK1-NEXT:    store ptr [[TMP11]], ptr [[TMP10]], align 8
 201 // CHECK1-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 2
 202 // CHECK1-NEXT:    store ptr [[SVAR]], ptr [[TMP12]], align 8
 203 // CHECK1-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 3
 204 // CHECK1-NEXT:    store ptr [[SFVAR]], ptr [[TMP13]], align 8
 205 // CHECK1-NEXT:    call void @"_ZZZ4mainENK3$_0clEvENKUlvE_clEv"(ptr noundef nonnull align 8 dereferenceable(32) [[REF_TMP]])
 206 // CHECK1-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]
 207 // CHECK1:       omp.body.continue:
 208 // CHECK1-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]
 209 // CHECK1:       omp.inner.for.inc:
 210 // CHECK1-NEXT:    [[TMP14:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 211 // CHECK1-NEXT:    [[ADD4:%.*]] = add nsw i32 [[TMP14]], 1
 212 // CHECK1-NEXT:    store i32 [[ADD4]], ptr [[DOTOMP_IV]], align 4
 213 // CHECK1-NEXT:    br label [[OMP_INNER_FOR_COND]]
 214 // CHECK1:       omp.inner.for.end:
 215 // CHECK1-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]
 216 // CHECK1:       omp.loop.exit:
 217 // CHECK1-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP1]])
 218 // CHECK1-NEXT:    ret void
 219 //
 220 //
 221 // CHECK3-LABEL: define {{[^@]+}}@main
 222 // CHECK3-SAME: () #[[ATTR0:[0-9]+]] {
 223 // CHECK3-NEXT:  entry:
 224 // CHECK3-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4
 225 // CHECK3-NEXT:    [[G:%.*]] = alloca double, align 8
 226 // CHECK3-NEXT:    [[G1:%.*]] = alloca ptr, align 4
 227 // CHECK3-NEXT:    [[REF_TMP:%.*]] = alloca [[CLASS_ANON:%.*]], align 1
 228 // CHECK3-NEXT:    store i32 0, ptr [[RETVAL]], align 4
 229 // CHECK3-NEXT:    store ptr [[G]], ptr [[G1]], align 4
 230 // CHECK3-NEXT:    call void @"_ZZ4mainENK3$_0clEv"(ptr noundef nonnull align 1 dereferenceable(1) [[REF_TMP]])
 231 // CHECK3-NEXT:    ret i32 0
 232 //
 233 //
 234 // CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68
 235 // CHECK3-SAME: () #[[ATTR2:[0-9]+]] {
 236 // CHECK3-NEXT:  entry:
 237 // CHECK3-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB2:[0-9]+]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined)
 238 // CHECK3-NEXT:    ret void
 239 //
 240 //
 241 // CHECK3-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l68.omp_outlined
 242 // CHECK3-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR2]] {
 243 // CHECK3-NEXT:  entry:
 244 // CHECK3-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4
 245 // CHECK3-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4
 246 // CHECK3-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4
 247 // CHECK3-NEXT:    [[TMP:%.*]] = alloca i32, align 4
 248 // CHECK3-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 4
 249 // CHECK3-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4
 250 // CHECK3-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4
 251 // CHECK3-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
 252 // CHECK3-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
 253 // CHECK3-NEXT:    [[G:%.*]] = alloca double, align 8
 254 // CHECK3-NEXT:    [[G1:%.*]] = alloca double, align 8
 255 // CHECK3-NEXT:    [[_TMP2:%.*]] = alloca ptr, align 4
 256 // CHECK3-NEXT:    [[SVAR:%.*]] = alloca i32, align 4
 257 // CHECK3-NEXT:    [[SFVAR:%.*]] = alloca float, align 4
 258 // CHECK3-NEXT:    [[I:%.*]] = alloca i32, align 4
 259 // CHECK3-NEXT:    [[REF_TMP:%.*]] = alloca [[CLASS_ANON_0:%.*]], align 4
 260 // CHECK3-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4
 261 // CHECK3-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4
 262 // CHECK3-NEXT:    store ptr undef, ptr [[_TMP1]], align 4
 263 // CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4
 264 // CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4
 265 // CHECK3-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4
 266 // CHECK3-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4
 267 // CHECK3-NEXT:    store ptr [[G1]], ptr [[_TMP2]], align 4
 268 // CHECK3-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4
 269 // CHECK3-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4
 270 // CHECK3-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
 271 // CHECK3-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 272 // CHECK3-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 1
 273 // CHECK3-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]
 274 // CHECK3:       cond.true:
 275 // CHECK3-NEXT:    br label [[COND_END:%.*]]
 276 // CHECK3:       cond.false:
 277 // CHECK3-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 278 // CHECK3-NEXT:    br label [[COND_END]]
 279 // CHECK3:       cond.end:
 280 // CHECK3-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]
 281 // CHECK3-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4
 282 // CHECK3-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4
 283 // CHECK3-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4
 284 // CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]
 285 // CHECK3:       omp.inner.for.cond:
 286 // CHECK3-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 287 // CHECK3-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 288 // CHECK3-NEXT:    [[CMP3:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]
 289 // CHECK3-NEXT:    br i1 [[CMP3]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]
 290 // CHECK3:       omp.inner.for.body:
 291 // CHECK3-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 292 // CHECK3-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP7]], 1
 293 // CHECK3-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]
 294 // CHECK3-NEXT:    store i32 [[ADD]], ptr [[I]], align 4
 295 // CHECK3-NEXT:    store double 1.000000e+00, ptr [[G]], align 8
 296 // CHECK3-NEXT:    [[TMP8:%.*]] = load ptr, ptr [[_TMP2]], align 4
 297 // CHECK3-NEXT:    store volatile double 1.000000e+00, ptr [[TMP8]], align 4
 298 // CHECK3-NEXT:    store i32 3, ptr [[SVAR]], align 4
 299 // CHECK3-NEXT:    store float 4.000000e+00, ptr [[SFVAR]], align 4
 300 // CHECK3-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 0
 301 // CHECK3-NEXT:    store ptr [[G]], ptr [[TMP9]], align 4
 302 // CHECK3-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 1
 303 // CHECK3-NEXT:    [[TMP11:%.*]] = load ptr, ptr [[_TMP2]], align 4
 304 // CHECK3-NEXT:    store ptr [[TMP11]], ptr [[TMP10]], align 4
 305 // CHECK3-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 2
 306 // CHECK3-NEXT:    store ptr [[SVAR]], ptr [[TMP12]], align 4
 307 // CHECK3-NEXT:    [[TMP13:%.*]] = getelementptr inbounds nuw [[CLASS_ANON_0]], ptr [[REF_TMP]], i32 0, i32 3
 308 // CHECK3-NEXT:    store ptr [[SFVAR]], ptr [[TMP13]], align 4
 309 // CHECK3-NEXT:    call void @"_ZZZ4mainENK3$_0clEvENKUlvE_clEv"(ptr noundef nonnull align 4 dereferenceable(16) [[REF_TMP]])
 310 // CHECK3-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]
 311 // CHECK3:       omp.body.continue:
 312 // CHECK3-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]
 313 // CHECK3:       omp.inner.for.inc:
 314 // CHECK3-NEXT:    [[TMP14:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 315 // CHECK3-NEXT:    [[ADD4:%.*]] = add nsw i32 [[TMP14]], 1
 316 // CHECK3-NEXT:    store i32 [[ADD4]], ptr [[DOTOMP_IV]], align 4
 317 // CHECK3-NEXT:    br label [[OMP_INNER_FOR_COND]]
 318 // CHECK3:       omp.inner.for.end:
 319 // CHECK3-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]
 320 // CHECK3:       omp.loop.exit:
 321 // CHECK3-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP1]])
 322 // CHECK3-NEXT:    ret void
 323 //
 324 //
 325 // CHECK9-LABEL: define {{[^@]+}}@main
 326 // CHECK9-SAME: () #[[ATTR0:[0-9]+]] {
 327 // CHECK9-NEXT:  entry:
 328 // CHECK9-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4
 329 // CHECK9-NEXT:    [[G:%.*]] = alloca double, align 8
 330 // CHECK9-NEXT:    [[G1:%.*]] = alloca ptr, align 8
 331 // CHECK9-NEXT:    [[TEST:%.*]] = alloca [[STRUCT_S:%.*]], align 4
 332 // CHECK9-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4
 333 // CHECK9-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4
 334 // CHECK9-NEXT:    [[S_ARR:%.*]] = alloca [2 x %struct.S], align 4
 335 // CHECK9-NEXT:    [[VAR:%.*]] = alloca ptr, align 8
 336 // CHECK9-NEXT:    [[TMP:%.*]] = alloca i32, align 4
 337 // CHECK9-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 8
 338 // CHECK9-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
 339 // CHECK9-NEXT:    [[I:%.*]] = alloca i32, align 4
 340 // CHECK9-NEXT:    [[_TMP2:%.*]] = alloca i32, align 4
 341 // CHECK9-NEXT:    [[KERNEL_ARGS3:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 8
 342 // CHECK9-NEXT:    store i32 0, ptr [[RETVAL]], align 4
 343 // CHECK9-NEXT:    store ptr [[G]], ptr [[G1]], align 8
 344 // CHECK9-NEXT:    call void @_ZN1SIfEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[TEST]])
 345 // CHECK9-NEXT:    store i32 0, ptr [[T_VAR]], align 4
 346 // CHECK9-NEXT:    call void @llvm.memcpy.p0.p0.i64(ptr align 4 [[VEC]], ptr align 4 @__const.main.vec, i64 8, i1 false)
 347 // CHECK9-NEXT:    call void @_ZN1SIfEC1Ef(ptr noundef nonnull align 4 dereferenceable(4) [[S_ARR]], float noundef 1.000000e+00)
 348 // CHECK9-NEXT:    [[ARRAYINIT_ELEMENT:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[S_ARR]], i64 1
 349 // CHECK9-NEXT:    call void @_ZN1SIfEC1Ef(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYINIT_ELEMENT]], float noundef 2.000000e+00)
 350 // CHECK9-NEXT:    store ptr [[TEST]], ptr [[VAR]], align 8
 351 // CHECK9-NEXT:    store ptr undef, ptr [[_TMP1]], align 8
 352 // CHECK9-NEXT:    [[TMP0:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0
 353 // CHECK9-NEXT:    store i32 3, ptr [[TMP0]], align 4
 354 // CHECK9-NEXT:    [[TMP1:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1
 355 // CHECK9-NEXT:    store i32 0, ptr [[TMP1]], align 4
 356 // CHECK9-NEXT:    [[TMP2:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2
 357 // CHECK9-NEXT:    store ptr null, ptr [[TMP2]], align 8
 358 // CHECK9-NEXT:    [[TMP3:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3
 359 // CHECK9-NEXT:    store ptr null, ptr [[TMP3]], align 8
 360 // CHECK9-NEXT:    [[TMP4:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4
 361 // CHECK9-NEXT:    store ptr null, ptr [[TMP4]], align 8
 362 // CHECK9-NEXT:    [[TMP5:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5
 363 // CHECK9-NEXT:    store ptr null, ptr [[TMP5]], align 8
 364 // CHECK9-NEXT:    [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6
 365 // CHECK9-NEXT:    store ptr null, ptr [[TMP6]], align 8
 366 // CHECK9-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7
 367 // CHECK9-NEXT:    store ptr null, ptr [[TMP7]], align 8
 368 // CHECK9-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8
 369 // CHECK9-NEXT:    store i64 2, ptr [[TMP8]], align 8
 370 // CHECK9-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9
 371 // CHECK9-NEXT:    store i64 0, ptr [[TMP9]], align 8
 372 // CHECK9-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10
 373 // CHECK9-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP10]], align 4
 374 // CHECK9-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11
 375 // CHECK9-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP11]], align 4
 376 // CHECK9-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12
 377 // CHECK9-NEXT:    store i32 0, ptr [[TMP12]], align 4
 378 // CHECK9-NEXT:    [[TMP13:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB2:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93.region_id, ptr [[KERNEL_ARGS]])
 379 // CHECK9-NEXT:    [[TMP14:%.*]] = icmp ne i32 [[TMP13]], 0
 380 // CHECK9-NEXT:    br i1 [[TMP14]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
 381 // CHECK9:       omp_offload.failed:
 382 // CHECK9-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93() #[[ATTR4:[0-9]+]]
 383 // CHECK9-NEXT:    br label [[OMP_OFFLOAD_CONT]]
 384 // CHECK9:       omp_offload.cont:
 385 // CHECK9-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 0
 386 // CHECK9-NEXT:    store i32 3, ptr [[TMP15]], align 4
 387 // CHECK9-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 1
 388 // CHECK9-NEXT:    store i32 0, ptr [[TMP16]], align 4
 389 // CHECK9-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 2
 390 // CHECK9-NEXT:    store ptr null, ptr [[TMP17]], align 8
 391 // CHECK9-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 3
 392 // CHECK9-NEXT:    store ptr null, ptr [[TMP18]], align 8
 393 // CHECK9-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 4
 394 // CHECK9-NEXT:    store ptr null, ptr [[TMP19]], align 8
 395 // CHECK9-NEXT:    [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 5
 396 // CHECK9-NEXT:    store ptr null, ptr [[TMP20]], align 8
 397 // CHECK9-NEXT:    [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 6
 398 // CHECK9-NEXT:    store ptr null, ptr [[TMP21]], align 8
 399 // CHECK9-NEXT:    [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 7
 400 // CHECK9-NEXT:    store ptr null, ptr [[TMP22]], align 8
 401 // CHECK9-NEXT:    [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 8
 402 // CHECK9-NEXT:    store i64 2, ptr [[TMP23]], align 8
 403 // CHECK9-NEXT:    [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 9
 404 // CHECK9-NEXT:    store i64 0, ptr [[TMP24]], align 8
 405 // CHECK9-NEXT:    [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 10
 406 // CHECK9-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP25]], align 4
 407 // CHECK9-NEXT:    [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 11
 408 // CHECK9-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP26]], align 4
 409 // CHECK9-NEXT:    [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 12
 410 // CHECK9-NEXT:    store i32 0, ptr [[TMP27]], align 4
 411 // CHECK9-NEXT:    [[TMP28:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB2]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102.region_id, ptr [[KERNEL_ARGS3]])
 412 // CHECK9-NEXT:    [[TMP29:%.*]] = icmp ne i32 [[TMP28]], 0
 413 // CHECK9-NEXT:    br i1 [[TMP29]], label [[OMP_OFFLOAD_FAILED4:%.*]], label [[OMP_OFFLOAD_CONT5:%.*]]
 414 // CHECK9:       omp_offload.failed4:
 415 // CHECK9-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102() #[[ATTR4]]
 416 // CHECK9-NEXT:    br label [[OMP_OFFLOAD_CONT5]]
 417 // CHECK9:       omp_offload.cont5:
 418 // CHECK9-NEXT:    [[CALL:%.*]] = call noundef signext i32 @_Z5tmainIiET_v()
 419 // CHECK9-NEXT:    store i32 [[CALL]], ptr [[RETVAL]], align 4
 420 // CHECK9-NEXT:    [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i32 0, i32 0
 421 // CHECK9-NEXT:    [[TMP30:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN]], i64 2
 422 // CHECK9-NEXT:    br label [[ARRAYDESTROY_BODY:%.*]]
 423 // CHECK9:       arraydestroy.body:
 424 // CHECK9-NEXT:    [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP30]], [[OMP_OFFLOAD_CONT5]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]
 425 // CHECK9-NEXT:    [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i64 -1
 426 // CHECK9-NEXT:    call void @_ZN1SIfED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]
 427 // CHECK9-NEXT:    [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN]]
 428 // CHECK9-NEXT:    br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE6:%.*]], label [[ARRAYDESTROY_BODY]]
 429 // CHECK9:       arraydestroy.done6:
 430 // CHECK9-NEXT:    call void @_ZN1SIfED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[TEST]]) #[[ATTR4]]
 431 // CHECK9-NEXT:    [[TMP31:%.*]] = load i32, ptr [[RETVAL]], align 4
 432 // CHECK9-NEXT:    ret i32 [[TMP31]]
 433 //
 434 //
 435 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC1Ev
 436 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1:[0-9]+]] comdat {
 437 // CHECK9-NEXT:  entry:
 438 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 439 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 440 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 441 // CHECK9-NEXT:    call void @_ZN1SIfEC2Ev(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]])
 442 // CHECK9-NEXT:    ret void
 443 //
 444 //
 445 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC1Ef
 446 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]], float noundef [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 447 // CHECK9-NEXT:  entry:
 448 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 449 // CHECK9-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
 450 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 451 // CHECK9-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
 452 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 453 // CHECK9-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
 454 // CHECK9-NEXT:    call void @_ZN1SIfEC2Ef(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]], float noundef [[TMP0]])
 455 // CHECK9-NEXT:    ret void
 456 //
 457 //
 458 // CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93
 459 // CHECK9-SAME: () #[[ATTR3:[0-9]+]] {
 460 // CHECK9-NEXT:  entry:
 461 // CHECK9-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB2]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93.omp_outlined)
 462 // CHECK9-NEXT:    ret void
 463 //
 464 //
 465 // CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93.omp_outlined
 466 // CHECK9-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR3]] {
 467 // CHECK9-NEXT:  entry:
 468 // CHECK9-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8
 469 // CHECK9-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8
 470 // CHECK9-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4
 471 // CHECK9-NEXT:    [[TMP:%.*]] = alloca i32, align 4
 472 // CHECK9-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 8
 473 // CHECK9-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4
 474 // CHECK9-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4
 475 // CHECK9-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
 476 // CHECK9-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
 477 // CHECK9-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4
 478 // CHECK9-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4
 479 // CHECK9-NEXT:    [[S_ARR:%.*]] = alloca [2 x %struct.S], align 4
 480 // CHECK9-NEXT:    [[VAR:%.*]] = alloca [[STRUCT_S:%.*]], align 4
 481 // CHECK9-NEXT:    [[_TMP2:%.*]] = alloca ptr, align 8
 482 // CHECK9-NEXT:    [[SVAR:%.*]] = alloca i32, align 4
 483 // CHECK9-NEXT:    [[I:%.*]] = alloca i32, align 4
 484 // CHECK9-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8
 485 // CHECK9-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8
 486 // CHECK9-NEXT:    store ptr undef, ptr [[_TMP1]], align 8
 487 // CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4
 488 // CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4
 489 // CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4
 490 // CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4
 491 // CHECK9-NEXT:    [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i32 0, i32 0
 492 // CHECK9-NEXT:    [[ARRAYCTOR_END:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN]], i64 2
 493 // CHECK9-NEXT:    br label [[ARRAYCTOR_LOOP:%.*]]
 494 // CHECK9:       arrayctor.loop:
 495 // CHECK9-NEXT:    [[ARRAYCTOR_CUR:%.*]] = phi ptr [ [[ARRAY_BEGIN]], [[ENTRY:%.*]] ], [ [[ARRAYCTOR_NEXT:%.*]], [[ARRAYCTOR_LOOP]] ]
 496 // CHECK9-NEXT:    call void @_ZN1SIfEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYCTOR_CUR]])
 497 // CHECK9-NEXT:    [[ARRAYCTOR_NEXT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYCTOR_CUR]], i64 1
 498 // CHECK9-NEXT:    [[ARRAYCTOR_DONE:%.*]] = icmp eq ptr [[ARRAYCTOR_NEXT]], [[ARRAYCTOR_END]]
 499 // CHECK9-NEXT:    br i1 [[ARRAYCTOR_DONE]], label [[ARRAYCTOR_CONT:%.*]], label [[ARRAYCTOR_LOOP]]
 500 // CHECK9:       arrayctor.cont:
 501 // CHECK9-NEXT:    call void @_ZN1SIfEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[VAR]])
 502 // CHECK9-NEXT:    store ptr [[VAR]], ptr [[_TMP2]], align 8
 503 // CHECK9-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8
 504 // CHECK9-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4
 505 // CHECK9-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
 506 // CHECK9-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 507 // CHECK9-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 1
 508 // CHECK9-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]
 509 // CHECK9:       cond.true:
 510 // CHECK9-NEXT:    br label [[COND_END:%.*]]
 511 // CHECK9:       cond.false:
 512 // CHECK9-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 513 // CHECK9-NEXT:    br label [[COND_END]]
 514 // CHECK9:       cond.end:
 515 // CHECK9-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]
 516 // CHECK9-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4
 517 // CHECK9-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4
 518 // CHECK9-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4
 519 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]
 520 // CHECK9:       omp.inner.for.cond:
 521 // CHECK9-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 522 // CHECK9-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 523 // CHECK9-NEXT:    [[CMP3:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]
 524 // CHECK9-NEXT:    br i1 [[CMP3]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_COND_CLEANUP:%.*]]
 525 // CHECK9:       omp.inner.for.cond.cleanup:
 526 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_END:%.*]]
 527 // CHECK9:       omp.inner.for.body:
 528 // CHECK9-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 529 // CHECK9-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP7]], 1
 530 // CHECK9-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]
 531 // CHECK9-NEXT:    store i32 [[ADD]], ptr [[I]], align 4
 532 // CHECK9-NEXT:    [[TMP8:%.*]] = load i32, ptr [[T_VAR]], align 4
 533 // CHECK9-NEXT:    [[TMP9:%.*]] = load i32, ptr [[I]], align 4
 534 // CHECK9-NEXT:    [[IDXPROM:%.*]] = sext i32 [[TMP9]] to i64
 535 // CHECK9-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[VEC]], i64 0, i64 [[IDXPROM]]
 536 // CHECK9-NEXT:    store i32 [[TMP8]], ptr [[ARRAYIDX]], align 4
 537 // CHECK9-NEXT:    [[TMP10:%.*]] = load ptr, ptr [[_TMP2]], align 8
 538 // CHECK9-NEXT:    [[TMP11:%.*]] = load i32, ptr [[I]], align 4
 539 // CHECK9-NEXT:    [[IDXPROM4:%.*]] = sext i32 [[TMP11]] to i64
 540 // CHECK9-NEXT:    [[ARRAYIDX5:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i64 0, i64 [[IDXPROM4]]
 541 // CHECK9-NEXT:    call void @llvm.memcpy.p0.p0.i64(ptr align 4 [[ARRAYIDX5]], ptr align 4 [[TMP10]], i64 4, i1 false)
 542 // CHECK9-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]
 543 // CHECK9:       omp.body.continue:
 544 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]
 545 // CHECK9:       omp.inner.for.inc:
 546 // CHECK9-NEXT:    [[TMP12:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 547 // CHECK9-NEXT:    [[ADD6:%.*]] = add nsw i32 [[TMP12]], 1
 548 // CHECK9-NEXT:    store i32 [[ADD6]], ptr [[DOTOMP_IV]], align 4
 549 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND]]
 550 // CHECK9:       omp.inner.for.end:
 551 // CHECK9-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]
 552 // CHECK9:       omp.loop.exit:
 553 // CHECK9-NEXT:    [[TMP13:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8
 554 // CHECK9-NEXT:    [[TMP14:%.*]] = load i32, ptr [[TMP13]], align 4
 555 // CHECK9-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP14]])
 556 // CHECK9-NEXT:    call void @_ZN1SIfED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[VAR]]) #[[ATTR4]]
 557 // CHECK9-NEXT:    [[ARRAY_BEGIN7:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i32 0, i32 0
 558 // CHECK9-NEXT:    [[TMP15:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN7]], i64 2
 559 // CHECK9-NEXT:    br label [[ARRAYDESTROY_BODY:%.*]]
 560 // CHECK9:       arraydestroy.body:
 561 // CHECK9-NEXT:    [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP15]], [[OMP_LOOP_EXIT]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]
 562 // CHECK9-NEXT:    [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i64 -1
 563 // CHECK9-NEXT:    call void @_ZN1SIfED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]
 564 // CHECK9-NEXT:    [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN7]]
 565 // CHECK9-NEXT:    br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]
 566 // CHECK9:       arraydestroy.done8:
 567 // CHECK9-NEXT:    ret void
 568 //
 569 //
 570 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfED1Ev
 571 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 572 // CHECK9-NEXT:  entry:
 573 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 574 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 575 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 576 // CHECK9-NEXT:    call void @_ZN1SIfED2Ev(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]]) #[[ATTR4]]
 577 // CHECK9-NEXT:    ret void
 578 //
 579 //
 580 // CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102
 581 // CHECK9-SAME: () #[[ATTR3]] {
 582 // CHECK9-NEXT:  entry:
 583 // CHECK9-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB2]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102.omp_outlined)
 584 // CHECK9-NEXT:    ret void
 585 //
 586 //
 587 // CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102.omp_outlined
 588 // CHECK9-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR3]] {
 589 // CHECK9-NEXT:  entry:
 590 // CHECK9-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8
 591 // CHECK9-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8
 592 // CHECK9-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4
 593 // CHECK9-NEXT:    [[TMP:%.*]] = alloca i32, align 4
 594 // CHECK9-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4
 595 // CHECK9-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4
 596 // CHECK9-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
 597 // CHECK9-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
 598 // CHECK9-NEXT:    [[I:%.*]] = alloca i32, align 4
 599 // CHECK9-NEXT:    [[I1:%.*]] = alloca i32, align 4
 600 // CHECK9-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8
 601 // CHECK9-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8
 602 // CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4
 603 // CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4
 604 // CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4
 605 // CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4
 606 // CHECK9-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8
 607 // CHECK9-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4
 608 // CHECK9-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
 609 // CHECK9-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 610 // CHECK9-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 1
 611 // CHECK9-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]
 612 // CHECK9:       cond.true:
 613 // CHECK9-NEXT:    br label [[COND_END:%.*]]
 614 // CHECK9:       cond.false:
 615 // CHECK9-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 616 // CHECK9-NEXT:    br label [[COND_END]]
 617 // CHECK9:       cond.end:
 618 // CHECK9-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]
 619 // CHECK9-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4
 620 // CHECK9-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4
 621 // CHECK9-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4
 622 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]
 623 // CHECK9:       omp.inner.for.cond:
 624 // CHECK9-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 625 // CHECK9-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 626 // CHECK9-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]
 627 // CHECK9-NEXT:    br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]
 628 // CHECK9:       omp.inner.for.body:
 629 // CHECK9-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 630 // CHECK9-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP7]], 1
 631 // CHECK9-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]
 632 // CHECK9-NEXT:    store i32 [[ADD]], ptr [[I]], align 4
 633 // CHECK9-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]
 634 // CHECK9:       omp.body.continue:
 635 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]
 636 // CHECK9:       omp.inner.for.inc:
 637 // CHECK9-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 638 // CHECK9-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP8]], 1
 639 // CHECK9-NEXT:    store i32 [[ADD3]], ptr [[DOTOMP_IV]], align 4
 640 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND]]
 641 // CHECK9:       omp.inner.for.end:
 642 // CHECK9-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]
 643 // CHECK9:       omp.loop.exit:
 644 // CHECK9-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP1]])
 645 // CHECK9-NEXT:    ret void
 646 //
 647 //
 648 // CHECK9-LABEL: define {{[^@]+}}@_Z5tmainIiET_v
 649 // CHECK9-SAME: () #[[ATTR1]] comdat {
 650 // CHECK9-NEXT:  entry:
 651 // CHECK9-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4
 652 // CHECK9-NEXT:    [[TEST:%.*]] = alloca [[STRUCT_S_0:%.*]], align 4
 653 // CHECK9-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4
 654 // CHECK9-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4
 655 // CHECK9-NEXT:    [[S_ARR:%.*]] = alloca [2 x %struct.S.0], align 4
 656 // CHECK9-NEXT:    [[VAR:%.*]] = alloca ptr, align 8
 657 // CHECK9-NEXT:    [[TMP:%.*]] = alloca i32, align 4
 658 // CHECK9-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 8
 659 // CHECK9-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
 660 // CHECK9-NEXT:    call void @_ZN1SIiEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[TEST]])
 661 // CHECK9-NEXT:    store i32 0, ptr [[T_VAR]], align 4
 662 // CHECK9-NEXT:    call void @llvm.memcpy.p0.p0.i64(ptr align 4 [[VEC]], ptr align 4 @__const._Z5tmainIiET_v.vec, i64 8, i1 false)
 663 // CHECK9-NEXT:    call void @_ZN1SIiEC1Ei(ptr noundef nonnull align 4 dereferenceable(4) [[S_ARR]], i32 noundef signext 1)
 664 // CHECK9-NEXT:    [[ARRAYINIT_ELEMENT:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[S_ARR]], i64 1
 665 // CHECK9-NEXT:    call void @_ZN1SIiEC1Ei(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYINIT_ELEMENT]], i32 noundef signext 2)
 666 // CHECK9-NEXT:    store ptr [[TEST]], ptr [[VAR]], align 8
 667 // CHECK9-NEXT:    store ptr undef, ptr [[_TMP1]], align 8
 668 // CHECK9-NEXT:    [[TMP0:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0
 669 // CHECK9-NEXT:    store i32 3, ptr [[TMP0]], align 4
 670 // CHECK9-NEXT:    [[TMP1:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1
 671 // CHECK9-NEXT:    store i32 0, ptr [[TMP1]], align 4
 672 // CHECK9-NEXT:    [[TMP2:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2
 673 // CHECK9-NEXT:    store ptr null, ptr [[TMP2]], align 8
 674 // CHECK9-NEXT:    [[TMP3:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3
 675 // CHECK9-NEXT:    store ptr null, ptr [[TMP3]], align 8
 676 // CHECK9-NEXT:    [[TMP4:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4
 677 // CHECK9-NEXT:    store ptr null, ptr [[TMP4]], align 8
 678 // CHECK9-NEXT:    [[TMP5:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5
 679 // CHECK9-NEXT:    store ptr null, ptr [[TMP5]], align 8
 680 // CHECK9-NEXT:    [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6
 681 // CHECK9-NEXT:    store ptr null, ptr [[TMP6]], align 8
 682 // CHECK9-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7
 683 // CHECK9-NEXT:    store ptr null, ptr [[TMP7]], align 8
 684 // CHECK9-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8
 685 // CHECK9-NEXT:    store i64 2, ptr [[TMP8]], align 8
 686 // CHECK9-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9
 687 // CHECK9-NEXT:    store i64 0, ptr [[TMP9]], align 8
 688 // CHECK9-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10
 689 // CHECK9-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP10]], align 4
 690 // CHECK9-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11
 691 // CHECK9-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP11]], align 4
 692 // CHECK9-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12
 693 // CHECK9-NEXT:    store i32 0, ptr [[TMP12]], align 4
 694 // CHECK9-NEXT:    [[TMP13:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB2]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49.region_id, ptr [[KERNEL_ARGS]])
 695 // CHECK9-NEXT:    [[TMP14:%.*]] = icmp ne i32 [[TMP13]], 0
 696 // CHECK9-NEXT:    br i1 [[TMP14]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
 697 // CHECK9:       omp_offload.failed:
 698 // CHECK9-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49() #[[ATTR4]]
 699 // CHECK9-NEXT:    br label [[OMP_OFFLOAD_CONT]]
 700 // CHECK9:       omp_offload.cont:
 701 // CHECK9-NEXT:    store i32 0, ptr [[RETVAL]], align 4
 702 // CHECK9-NEXT:    [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i32 0, i32 0
 703 // CHECK9-NEXT:    [[TMP15:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN]], i64 2
 704 // CHECK9-NEXT:    br label [[ARRAYDESTROY_BODY:%.*]]
 705 // CHECK9:       arraydestroy.body:
 706 // CHECK9-NEXT:    [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP15]], [[OMP_OFFLOAD_CONT]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]
 707 // CHECK9-NEXT:    [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i64 -1
 708 // CHECK9-NEXT:    call void @_ZN1SIiED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]
 709 // CHECK9-NEXT:    [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN]]
 710 // CHECK9-NEXT:    br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE2:%.*]], label [[ARRAYDESTROY_BODY]]
 711 // CHECK9:       arraydestroy.done2:
 712 // CHECK9-NEXT:    call void @_ZN1SIiED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[TEST]]) #[[ATTR4]]
 713 // CHECK9-NEXT:    [[TMP16:%.*]] = load i32, ptr [[RETVAL]], align 4
 714 // CHECK9-NEXT:    ret i32 [[TMP16]]
 715 //
 716 //
 717 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC2Ev
 718 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 719 // CHECK9-NEXT:  entry:
 720 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 721 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 722 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 723 // CHECK9-NEXT:    [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0
 724 // CHECK9-NEXT:    store float 0.000000e+00, ptr [[F]], align 4
 725 // CHECK9-NEXT:    ret void
 726 //
 727 //
 728 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfEC2Ef
 729 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]], float noundef [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 730 // CHECK9-NEXT:  entry:
 731 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 732 // CHECK9-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
 733 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 734 // CHECK9-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
 735 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 736 // CHECK9-NEXT:    [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0
 737 // CHECK9-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
 738 // CHECK9-NEXT:    store float [[TMP0]], ptr [[F]], align 4
 739 // CHECK9-NEXT:    ret void
 740 //
 741 //
 742 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIfED2Ev
 743 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 744 // CHECK9-NEXT:  entry:
 745 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 746 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 747 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 748 // CHECK9-NEXT:    ret void
 749 //
 750 //
 751 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC1Ev
 752 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 753 // CHECK9-NEXT:  entry:
 754 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 755 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 756 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 757 // CHECK9-NEXT:    call void @_ZN1SIiEC2Ev(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]])
 758 // CHECK9-NEXT:    ret void
 759 //
 760 //
 761 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC1Ei
 762 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]], i32 noundef signext [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 763 // CHECK9-NEXT:  entry:
 764 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 765 // CHECK9-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4
 766 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 767 // CHECK9-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 4
 768 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 769 // CHECK9-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4
 770 // CHECK9-NEXT:    call void @_ZN1SIiEC2Ei(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]], i32 noundef signext [[TMP0]])
 771 // CHECK9-NEXT:    ret void
 772 //
 773 //
 774 // CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49
 775 // CHECK9-SAME: () #[[ATTR3]] {
 776 // CHECK9-NEXT:  entry:
 777 // CHECK9-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB2]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49.omp_outlined)
 778 // CHECK9-NEXT:    ret void
 779 //
 780 //
 781 // CHECK9-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49.omp_outlined
 782 // CHECK9-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR3]] {
 783 // CHECK9-NEXT:  entry:
 784 // CHECK9-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8
 785 // CHECK9-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8
 786 // CHECK9-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4
 787 // CHECK9-NEXT:    [[TMP:%.*]] = alloca i32, align 4
 788 // CHECK9-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 8
 789 // CHECK9-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4
 790 // CHECK9-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4
 791 // CHECK9-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
 792 // CHECK9-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
 793 // CHECK9-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4
 794 // CHECK9-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4
 795 // CHECK9-NEXT:    [[S_ARR:%.*]] = alloca [2 x %struct.S.0], align 4
 796 // CHECK9-NEXT:    [[VAR:%.*]] = alloca [[STRUCT_S_0:%.*]], align 4
 797 // CHECK9-NEXT:    [[_TMP2:%.*]] = alloca ptr, align 8
 798 // CHECK9-NEXT:    [[I:%.*]] = alloca i32, align 4
 799 // CHECK9-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 8
 800 // CHECK9-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 8
 801 // CHECK9-NEXT:    store ptr undef, ptr [[_TMP1]], align 8
 802 // CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4
 803 // CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4
 804 // CHECK9-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4
 805 // CHECK9-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4
 806 // CHECK9-NEXT:    [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i32 0, i32 0
 807 // CHECK9-NEXT:    [[ARRAYCTOR_END:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN]], i64 2
 808 // CHECK9-NEXT:    br label [[ARRAYCTOR_LOOP:%.*]]
 809 // CHECK9:       arrayctor.loop:
 810 // CHECK9-NEXT:    [[ARRAYCTOR_CUR:%.*]] = phi ptr [ [[ARRAY_BEGIN]], [[ENTRY:%.*]] ], [ [[ARRAYCTOR_NEXT:%.*]], [[ARRAYCTOR_LOOP]] ]
 811 // CHECK9-NEXT:    call void @_ZN1SIiEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYCTOR_CUR]])
 812 // CHECK9-NEXT:    [[ARRAYCTOR_NEXT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYCTOR_CUR]], i64 1
 813 // CHECK9-NEXT:    [[ARRAYCTOR_DONE:%.*]] = icmp eq ptr [[ARRAYCTOR_NEXT]], [[ARRAYCTOR_END]]
 814 // CHECK9-NEXT:    br i1 [[ARRAYCTOR_DONE]], label [[ARRAYCTOR_CONT:%.*]], label [[ARRAYCTOR_LOOP]]
 815 // CHECK9:       arrayctor.cont:
 816 // CHECK9-NEXT:    call void @_ZN1SIiEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[VAR]])
 817 // CHECK9-NEXT:    store ptr [[VAR]], ptr [[_TMP2]], align 8
 818 // CHECK9-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8
 819 // CHECK9-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4
 820 // CHECK9-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
 821 // CHECK9-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 822 // CHECK9-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 1
 823 // CHECK9-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]
 824 // CHECK9:       cond.true:
 825 // CHECK9-NEXT:    br label [[COND_END:%.*]]
 826 // CHECK9:       cond.false:
 827 // CHECK9-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 828 // CHECK9-NEXT:    br label [[COND_END]]
 829 // CHECK9:       cond.end:
 830 // CHECK9-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]
 831 // CHECK9-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4
 832 // CHECK9-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4
 833 // CHECK9-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4
 834 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]
 835 // CHECK9:       omp.inner.for.cond:
 836 // CHECK9-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 837 // CHECK9-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
 838 // CHECK9-NEXT:    [[CMP3:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]
 839 // CHECK9-NEXT:    br i1 [[CMP3]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_COND_CLEANUP:%.*]]
 840 // CHECK9:       omp.inner.for.cond.cleanup:
 841 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_END:%.*]]
 842 // CHECK9:       omp.inner.for.body:
 843 // CHECK9-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 844 // CHECK9-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP7]], 1
 845 // CHECK9-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]
 846 // CHECK9-NEXT:    store i32 [[ADD]], ptr [[I]], align 4
 847 // CHECK9-NEXT:    [[TMP8:%.*]] = load i32, ptr [[T_VAR]], align 4
 848 // CHECK9-NEXT:    [[TMP9:%.*]] = load i32, ptr [[I]], align 4
 849 // CHECK9-NEXT:    [[IDXPROM:%.*]] = sext i32 [[TMP9]] to i64
 850 // CHECK9-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[VEC]], i64 0, i64 [[IDXPROM]]
 851 // CHECK9-NEXT:    store i32 [[TMP8]], ptr [[ARRAYIDX]], align 4
 852 // CHECK9-NEXT:    [[TMP10:%.*]] = load ptr, ptr [[_TMP2]], align 8
 853 // CHECK9-NEXT:    [[TMP11:%.*]] = load i32, ptr [[I]], align 4
 854 // CHECK9-NEXT:    [[IDXPROM4:%.*]] = sext i32 [[TMP11]] to i64
 855 // CHECK9-NEXT:    [[ARRAYIDX5:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i64 0, i64 [[IDXPROM4]]
 856 // CHECK9-NEXT:    call void @llvm.memcpy.p0.p0.i64(ptr align 4 [[ARRAYIDX5]], ptr align 4 [[TMP10]], i64 4, i1 false)
 857 // CHECK9-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]
 858 // CHECK9:       omp.body.continue:
 859 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]
 860 // CHECK9:       omp.inner.for.inc:
 861 // CHECK9-NEXT:    [[TMP12:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
 862 // CHECK9-NEXT:    [[ADD6:%.*]] = add nsw i32 [[TMP12]], 1
 863 // CHECK9-NEXT:    store i32 [[ADD6]], ptr [[DOTOMP_IV]], align 4
 864 // CHECK9-NEXT:    br label [[OMP_INNER_FOR_COND]]
 865 // CHECK9:       omp.inner.for.end:
 866 // CHECK9-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]
 867 // CHECK9:       omp.loop.exit:
 868 // CHECK9-NEXT:    [[TMP13:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 8
 869 // CHECK9-NEXT:    [[TMP14:%.*]] = load i32, ptr [[TMP13]], align 4
 870 // CHECK9-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP14]])
 871 // CHECK9-NEXT:    call void @_ZN1SIiED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[VAR]]) #[[ATTR4]]
 872 // CHECK9-NEXT:    [[ARRAY_BEGIN7:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i32 0, i32 0
 873 // CHECK9-NEXT:    [[TMP15:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN7]], i64 2
 874 // CHECK9-NEXT:    br label [[ARRAYDESTROY_BODY:%.*]]
 875 // CHECK9:       arraydestroy.body:
 876 // CHECK9-NEXT:    [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP15]], [[OMP_LOOP_EXIT]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]
 877 // CHECK9-NEXT:    [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i64 -1
 878 // CHECK9-NEXT:    call void @_ZN1SIiED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]
 879 // CHECK9-NEXT:    [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN7]]
 880 // CHECK9-NEXT:    br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE8:%.*]], label [[ARRAYDESTROY_BODY]]
 881 // CHECK9:       arraydestroy.done8:
 882 // CHECK9-NEXT:    ret void
 883 //
 884 //
 885 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiED1Ev
 886 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 887 // CHECK9-NEXT:  entry:
 888 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 889 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 890 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 891 // CHECK9-NEXT:    call void @_ZN1SIiED2Ev(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]]) #[[ATTR4]]
 892 // CHECK9-NEXT:    ret void
 893 //
 894 //
 895 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC2Ev
 896 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 897 // CHECK9-NEXT:  entry:
 898 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 899 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 900 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 901 // CHECK9-NEXT:    [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 0
 902 // CHECK9-NEXT:    store i32 0, ptr [[F]], align 4
 903 // CHECK9-NEXT:    ret void
 904 //
 905 //
 906 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiEC2Ei
 907 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]], i32 noundef signext [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 908 // CHECK9-NEXT:  entry:
 909 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 910 // CHECK9-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4
 911 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 912 // CHECK9-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 4
 913 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 914 // CHECK9-NEXT:    [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 0
 915 // CHECK9-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4
 916 // CHECK9-NEXT:    store i32 [[TMP0]], ptr [[F]], align 4
 917 // CHECK9-NEXT:    ret void
 918 //
 919 //
 920 // CHECK9-LABEL: define {{[^@]+}}@_ZN1SIiED2Ev
 921 // CHECK9-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat {
 922 // CHECK9-NEXT:  entry:
 923 // CHECK9-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
 924 // CHECK9-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
 925 // CHECK9-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
 926 // CHECK9-NEXT:    ret void
 927 //
 928 //
 929 // CHECK11-LABEL: define {{[^@]+}}@main
 930 // CHECK11-SAME: () #[[ATTR0:[0-9]+]] {
 931 // CHECK11-NEXT:  entry:
 932 // CHECK11-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4
 933 // CHECK11-NEXT:    [[G:%.*]] = alloca double, align 8
 934 // CHECK11-NEXT:    [[G1:%.*]] = alloca ptr, align 4
 935 // CHECK11-NEXT:    [[TEST:%.*]] = alloca [[STRUCT_S:%.*]], align 4
 936 // CHECK11-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4
 937 // CHECK11-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4
 938 // CHECK11-NEXT:    [[S_ARR:%.*]] = alloca [2 x %struct.S], align 4
 939 // CHECK11-NEXT:    [[VAR:%.*]] = alloca ptr, align 4
 940 // CHECK11-NEXT:    [[TMP:%.*]] = alloca i32, align 4
 941 // CHECK11-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 4
 942 // CHECK11-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
 943 // CHECK11-NEXT:    [[I:%.*]] = alloca i32, align 4
 944 // CHECK11-NEXT:    [[_TMP2:%.*]] = alloca i32, align 4
 945 // CHECK11-NEXT:    [[KERNEL_ARGS3:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 8
 946 // CHECK11-NEXT:    store i32 0, ptr [[RETVAL]], align 4
 947 // CHECK11-NEXT:    store ptr [[G]], ptr [[G1]], align 4
 948 // CHECK11-NEXT:    call void @_ZN1SIfEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[TEST]])
 949 // CHECK11-NEXT:    store i32 0, ptr [[T_VAR]], align 4
 950 // CHECK11-NEXT:    call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[VEC]], ptr align 4 @__const.main.vec, i32 8, i1 false)
 951 // CHECK11-NEXT:    call void @_ZN1SIfEC1Ef(ptr noundef nonnull align 4 dereferenceable(4) [[S_ARR]], float noundef 1.000000e+00)
 952 // CHECK11-NEXT:    [[ARRAYINIT_ELEMENT:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[S_ARR]], i32 1
 953 // CHECK11-NEXT:    call void @_ZN1SIfEC1Ef(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYINIT_ELEMENT]], float noundef 2.000000e+00)
 954 // CHECK11-NEXT:    store ptr [[TEST]], ptr [[VAR]], align 4
 955 // CHECK11-NEXT:    store ptr undef, ptr [[_TMP1]], align 4
 956 // CHECK11-NEXT:    [[TMP0:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0
 957 // CHECK11-NEXT:    store i32 3, ptr [[TMP0]], align 4
 958 // CHECK11-NEXT:    [[TMP1:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1
 959 // CHECK11-NEXT:    store i32 0, ptr [[TMP1]], align 4
 960 // CHECK11-NEXT:    [[TMP2:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2
 961 // CHECK11-NEXT:    store ptr null, ptr [[TMP2]], align 4
 962 // CHECK11-NEXT:    [[TMP3:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3
 963 // CHECK11-NEXT:    store ptr null, ptr [[TMP3]], align 4
 964 // CHECK11-NEXT:    [[TMP4:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4
 965 // CHECK11-NEXT:    store ptr null, ptr [[TMP4]], align 4
 966 // CHECK11-NEXT:    [[TMP5:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5
 967 // CHECK11-NEXT:    store ptr null, ptr [[TMP5]], align 4
 968 // CHECK11-NEXT:    [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6
 969 // CHECK11-NEXT:    store ptr null, ptr [[TMP6]], align 4
 970 // CHECK11-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7
 971 // CHECK11-NEXT:    store ptr null, ptr [[TMP7]], align 4
 972 // CHECK11-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8
 973 // CHECK11-NEXT:    store i64 2, ptr [[TMP8]], align 8
 974 // CHECK11-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9
 975 // CHECK11-NEXT:    store i64 0, ptr [[TMP9]], align 8
 976 // CHECK11-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10
 977 // CHECK11-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP10]], align 4
 978 // CHECK11-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11
 979 // CHECK11-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP11]], align 4
 980 // CHECK11-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12
 981 // CHECK11-NEXT:    store i32 0, ptr [[TMP12]], align 4
 982 // CHECK11-NEXT:    [[TMP13:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB2:[0-9]+]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93.region_id, ptr [[KERNEL_ARGS]])
 983 // CHECK11-NEXT:    [[TMP14:%.*]] = icmp ne i32 [[TMP13]], 0
 984 // CHECK11-NEXT:    br i1 [[TMP14]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
 985 // CHECK11:       omp_offload.failed:
 986 // CHECK11-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93() #[[ATTR4:[0-9]+]]
 987 // CHECK11-NEXT:    br label [[OMP_OFFLOAD_CONT]]
 988 // CHECK11:       omp_offload.cont:
 989 // CHECK11-NEXT:    [[TMP15:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 0
 990 // CHECK11-NEXT:    store i32 3, ptr [[TMP15]], align 4
 991 // CHECK11-NEXT:    [[TMP16:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 1
 992 // CHECK11-NEXT:    store i32 0, ptr [[TMP16]], align 4
 993 // CHECK11-NEXT:    [[TMP17:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 2
 994 // CHECK11-NEXT:    store ptr null, ptr [[TMP17]], align 4
 995 // CHECK11-NEXT:    [[TMP18:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 3
 996 // CHECK11-NEXT:    store ptr null, ptr [[TMP18]], align 4
 997 // CHECK11-NEXT:    [[TMP19:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 4
 998 // CHECK11-NEXT:    store ptr null, ptr [[TMP19]], align 4
 999 // CHECK11-NEXT:    [[TMP20:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 5
1000 // CHECK11-NEXT:    store ptr null, ptr [[TMP20]], align 4
1001 // CHECK11-NEXT:    [[TMP21:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 6
1002 // CHECK11-NEXT:    store ptr null, ptr [[TMP21]], align 4
1003 // CHECK11-NEXT:    [[TMP22:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 7
1004 // CHECK11-NEXT:    store ptr null, ptr [[TMP22]], align 4
1005 // CHECK11-NEXT:    [[TMP23:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 8
1006 // CHECK11-NEXT:    store i64 2, ptr [[TMP23]], align 8
1007 // CHECK11-NEXT:    [[TMP24:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 9
1008 // CHECK11-NEXT:    store i64 0, ptr [[TMP24]], align 8
1009 // CHECK11-NEXT:    [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 10
1010 // CHECK11-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP25]], align 4
1011 // CHECK11-NEXT:    [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 11
1012 // CHECK11-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP26]], align 4
1013 // CHECK11-NEXT:    [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS3]], i32 0, i32 12
1014 // CHECK11-NEXT:    store i32 0, ptr [[TMP27]], align 4
1015 // CHECK11-NEXT:    [[TMP28:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB2]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102.region_id, ptr [[KERNEL_ARGS3]])
1016 // CHECK11-NEXT:    [[TMP29:%.*]] = icmp ne i32 [[TMP28]], 0
1017 // CHECK11-NEXT:    br i1 [[TMP29]], label [[OMP_OFFLOAD_FAILED4:%.*]], label [[OMP_OFFLOAD_CONT5:%.*]]
1018 // CHECK11:       omp_offload.failed4:
1019 // CHECK11-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102() #[[ATTR4]]
1020 // CHECK11-NEXT:    br label [[OMP_OFFLOAD_CONT5]]
1021 // CHECK11:       omp_offload.cont5:
1022 // CHECK11-NEXT:    [[CALL:%.*]] = call noundef i32 @_Z5tmainIiET_v()
1023 // CHECK11-NEXT:    store i32 [[CALL]], ptr [[RETVAL]], align 4
1024 // CHECK11-NEXT:    [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i32 0, i32 0
1025 // CHECK11-NEXT:    [[TMP30:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN]], i32 2
1026 // CHECK11-NEXT:    br label [[ARRAYDESTROY_BODY:%.*]]
1027 // CHECK11:       arraydestroy.body:
1028 // CHECK11-NEXT:    [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP30]], [[OMP_OFFLOAD_CONT5]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]
1029 // CHECK11-NEXT:    [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i32 -1
1030 // CHECK11-NEXT:    call void @_ZN1SIfED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]
1031 // CHECK11-NEXT:    [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN]]
1032 // CHECK11-NEXT:    br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE6:%.*]], label [[ARRAYDESTROY_BODY]]
1033 // CHECK11:       arraydestroy.done6:
1034 // CHECK11-NEXT:    call void @_ZN1SIfED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[TEST]]) #[[ATTR4]]
1035 // CHECK11-NEXT:    [[TMP31:%.*]] = load i32, ptr [[RETVAL]], align 4
1036 // CHECK11-NEXT:    ret i32 [[TMP31]]
1037 //
1038 //
1039 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC1Ev
1040 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1:[0-9]+]] comdat align 2 {
1041 // CHECK11-NEXT:  entry:
1042 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1043 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1044 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1045 // CHECK11-NEXT:    call void @_ZN1SIfEC2Ev(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]])
1046 // CHECK11-NEXT:    ret void
1047 //
1048 //
1049 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC1Ef
1050 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]], float noundef [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1051 // CHECK11-NEXT:  entry:
1052 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1053 // CHECK11-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
1054 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1055 // CHECK11-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
1056 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1057 // CHECK11-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
1058 // CHECK11-NEXT:    call void @_ZN1SIfEC2Ef(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]], float noundef [[TMP0]])
1059 // CHECK11-NEXT:    ret void
1060 //
1061 //
1062 // CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93
1063 // CHECK11-SAME: () #[[ATTR3:[0-9]+]] {
1064 // CHECK11-NEXT:  entry:
1065 // CHECK11-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB2]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93.omp_outlined)
1066 // CHECK11-NEXT:    ret void
1067 //
1068 //
1069 // CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l93.omp_outlined
1070 // CHECK11-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR3]] {
1071 // CHECK11-NEXT:  entry:
1072 // CHECK11-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4
1073 // CHECK11-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4
1074 // CHECK11-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4
1075 // CHECK11-NEXT:    [[TMP:%.*]] = alloca i32, align 4
1076 // CHECK11-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 4
1077 // CHECK11-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4
1078 // CHECK11-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4
1079 // CHECK11-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
1080 // CHECK11-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
1081 // CHECK11-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4
1082 // CHECK11-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4
1083 // CHECK11-NEXT:    [[S_ARR:%.*]] = alloca [2 x %struct.S], align 4
1084 // CHECK11-NEXT:    [[VAR:%.*]] = alloca [[STRUCT_S:%.*]], align 4
1085 // CHECK11-NEXT:    [[_TMP2:%.*]] = alloca ptr, align 4
1086 // CHECK11-NEXT:    [[SVAR:%.*]] = alloca i32, align 4
1087 // CHECK11-NEXT:    [[I:%.*]] = alloca i32, align 4
1088 // CHECK11-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4
1089 // CHECK11-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4
1090 // CHECK11-NEXT:    store ptr undef, ptr [[_TMP1]], align 4
1091 // CHECK11-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4
1092 // CHECK11-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4
1093 // CHECK11-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4
1094 // CHECK11-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4
1095 // CHECK11-NEXT:    [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i32 0, i32 0
1096 // CHECK11-NEXT:    [[ARRAYCTOR_END:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN]], i32 2
1097 // CHECK11-NEXT:    br label [[ARRAYCTOR_LOOP:%.*]]
1098 // CHECK11:       arrayctor.loop:
1099 // CHECK11-NEXT:    [[ARRAYCTOR_CUR:%.*]] = phi ptr [ [[ARRAY_BEGIN]], [[ENTRY:%.*]] ], [ [[ARRAYCTOR_NEXT:%.*]], [[ARRAYCTOR_LOOP]] ]
1100 // CHECK11-NEXT:    call void @_ZN1SIfEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYCTOR_CUR]])
1101 // CHECK11-NEXT:    [[ARRAYCTOR_NEXT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYCTOR_CUR]], i32 1
1102 // CHECK11-NEXT:    [[ARRAYCTOR_DONE:%.*]] = icmp eq ptr [[ARRAYCTOR_NEXT]], [[ARRAYCTOR_END]]
1103 // CHECK11-NEXT:    br i1 [[ARRAYCTOR_DONE]], label [[ARRAYCTOR_CONT:%.*]], label [[ARRAYCTOR_LOOP]]
1104 // CHECK11:       arrayctor.cont:
1105 // CHECK11-NEXT:    call void @_ZN1SIfEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[VAR]])
1106 // CHECK11-NEXT:    store ptr [[VAR]], ptr [[_TMP2]], align 4
1107 // CHECK11-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4
1108 // CHECK11-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4
1109 // CHECK11-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1:[0-9]+]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
1110 // CHECK11-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
1111 // CHECK11-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 1
1112 // CHECK11-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]
1113 // CHECK11:       cond.true:
1114 // CHECK11-NEXT:    br label [[COND_END:%.*]]
1115 // CHECK11:       cond.false:
1116 // CHECK11-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
1117 // CHECK11-NEXT:    br label [[COND_END]]
1118 // CHECK11:       cond.end:
1119 // CHECK11-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]
1120 // CHECK11-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4
1121 // CHECK11-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4
1122 // CHECK11-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4
1123 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]
1124 // CHECK11:       omp.inner.for.cond:
1125 // CHECK11-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
1126 // CHECK11-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
1127 // CHECK11-NEXT:    [[CMP3:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]
1128 // CHECK11-NEXT:    br i1 [[CMP3]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_COND_CLEANUP:%.*]]
1129 // CHECK11:       omp.inner.for.cond.cleanup:
1130 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_END:%.*]]
1131 // CHECK11:       omp.inner.for.body:
1132 // CHECK11-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
1133 // CHECK11-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP7]], 1
1134 // CHECK11-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]
1135 // CHECK11-NEXT:    store i32 [[ADD]], ptr [[I]], align 4
1136 // CHECK11-NEXT:    [[TMP8:%.*]] = load i32, ptr [[T_VAR]], align 4
1137 // CHECK11-NEXT:    [[TMP9:%.*]] = load i32, ptr [[I]], align 4
1138 // CHECK11-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[VEC]], i32 0, i32 [[TMP9]]
1139 // CHECK11-NEXT:    store i32 [[TMP8]], ptr [[ARRAYIDX]], align 4
1140 // CHECK11-NEXT:    [[TMP10:%.*]] = load ptr, ptr [[_TMP2]], align 4
1141 // CHECK11-NEXT:    [[TMP11:%.*]] = load i32, ptr [[I]], align 4
1142 // CHECK11-NEXT:    [[ARRAYIDX4:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i32 0, i32 [[TMP11]]
1143 // CHECK11-NEXT:    call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[ARRAYIDX4]], ptr align 4 [[TMP10]], i32 4, i1 false)
1144 // CHECK11-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]
1145 // CHECK11:       omp.body.continue:
1146 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]
1147 // CHECK11:       omp.inner.for.inc:
1148 // CHECK11-NEXT:    [[TMP12:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
1149 // CHECK11-NEXT:    [[ADD5:%.*]] = add nsw i32 [[TMP12]], 1
1150 // CHECK11-NEXT:    store i32 [[ADD5]], ptr [[DOTOMP_IV]], align 4
1151 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_COND]]
1152 // CHECK11:       omp.inner.for.end:
1153 // CHECK11-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]
1154 // CHECK11:       omp.loop.exit:
1155 // CHECK11-NEXT:    [[TMP13:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4
1156 // CHECK11-NEXT:    [[TMP14:%.*]] = load i32, ptr [[TMP13]], align 4
1157 // CHECK11-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP14]])
1158 // CHECK11-NEXT:    call void @_ZN1SIfED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[VAR]]) #[[ATTR4]]
1159 // CHECK11-NEXT:    [[ARRAY_BEGIN6:%.*]] = getelementptr inbounds [2 x %struct.S], ptr [[S_ARR]], i32 0, i32 0
1160 // CHECK11-NEXT:    [[TMP15:%.*]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAY_BEGIN6]], i32 2
1161 // CHECK11-NEXT:    br label [[ARRAYDESTROY_BODY:%.*]]
1162 // CHECK11:       arraydestroy.body:
1163 // CHECK11-NEXT:    [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP15]], [[OMP_LOOP_EXIT]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]
1164 // CHECK11-NEXT:    [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i32 -1
1165 // CHECK11-NEXT:    call void @_ZN1SIfED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]
1166 // CHECK11-NEXT:    [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN6]]
1167 // CHECK11-NEXT:    br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE7:%.*]], label [[ARRAYDESTROY_BODY]]
1168 // CHECK11:       arraydestroy.done7:
1169 // CHECK11-NEXT:    ret void
1170 //
1171 //
1172 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfED1Ev
1173 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1174 // CHECK11-NEXT:  entry:
1175 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1176 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1177 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1178 // CHECK11-NEXT:    call void @_ZN1SIfED2Ev(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]]) #[[ATTR4]]
1179 // CHECK11-NEXT:    ret void
1180 //
1181 //
1182 // CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102
1183 // CHECK11-SAME: () #[[ATTR3]] {
1184 // CHECK11-NEXT:  entry:
1185 // CHECK11-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB2]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102.omp_outlined)
1186 // CHECK11-NEXT:    ret void
1187 //
1188 //
1189 // CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l102.omp_outlined
1190 // CHECK11-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR3]] {
1191 // CHECK11-NEXT:  entry:
1192 // CHECK11-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4
1193 // CHECK11-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4
1194 // CHECK11-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4
1195 // CHECK11-NEXT:    [[TMP:%.*]] = alloca i32, align 4
1196 // CHECK11-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4
1197 // CHECK11-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4
1198 // CHECK11-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
1199 // CHECK11-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
1200 // CHECK11-NEXT:    [[I:%.*]] = alloca i32, align 4
1201 // CHECK11-NEXT:    [[I1:%.*]] = alloca i32, align 4
1202 // CHECK11-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4
1203 // CHECK11-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4
1204 // CHECK11-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4
1205 // CHECK11-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4
1206 // CHECK11-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4
1207 // CHECK11-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4
1208 // CHECK11-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4
1209 // CHECK11-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4
1210 // CHECK11-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
1211 // CHECK11-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
1212 // CHECK11-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 1
1213 // CHECK11-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]
1214 // CHECK11:       cond.true:
1215 // CHECK11-NEXT:    br label [[COND_END:%.*]]
1216 // CHECK11:       cond.false:
1217 // CHECK11-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
1218 // CHECK11-NEXT:    br label [[COND_END]]
1219 // CHECK11:       cond.end:
1220 // CHECK11-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]
1221 // CHECK11-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4
1222 // CHECK11-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4
1223 // CHECK11-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4
1224 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]
1225 // CHECK11:       omp.inner.for.cond:
1226 // CHECK11-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
1227 // CHECK11-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
1228 // CHECK11-NEXT:    [[CMP2:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]
1229 // CHECK11-NEXT:    br i1 [[CMP2]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_END:%.*]]
1230 // CHECK11:       omp.inner.for.body:
1231 // CHECK11-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
1232 // CHECK11-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP7]], 1
1233 // CHECK11-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]
1234 // CHECK11-NEXT:    store i32 [[ADD]], ptr [[I]], align 4
1235 // CHECK11-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]
1236 // CHECK11:       omp.body.continue:
1237 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]
1238 // CHECK11:       omp.inner.for.inc:
1239 // CHECK11-NEXT:    [[TMP8:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
1240 // CHECK11-NEXT:    [[ADD3:%.*]] = add nsw i32 [[TMP8]], 1
1241 // CHECK11-NEXT:    store i32 [[ADD3]], ptr [[DOTOMP_IV]], align 4
1242 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_COND]]
1243 // CHECK11:       omp.inner.for.end:
1244 // CHECK11-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]
1245 // CHECK11:       omp.loop.exit:
1246 // CHECK11-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP1]])
1247 // CHECK11-NEXT:    ret void
1248 //
1249 //
1250 // CHECK11-LABEL: define {{[^@]+}}@_Z5tmainIiET_v
1251 // CHECK11-SAME: () #[[ATTR1]] comdat {
1252 // CHECK11-NEXT:  entry:
1253 // CHECK11-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4
1254 // CHECK11-NEXT:    [[TEST:%.*]] = alloca [[STRUCT_S_0:%.*]], align 4
1255 // CHECK11-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4
1256 // CHECK11-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4
1257 // CHECK11-NEXT:    [[S_ARR:%.*]] = alloca [2 x %struct.S.0], align 4
1258 // CHECK11-NEXT:    [[VAR:%.*]] = alloca ptr, align 4
1259 // CHECK11-NEXT:    [[TMP:%.*]] = alloca i32, align 4
1260 // CHECK11-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 4
1261 // CHECK11-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
1262 // CHECK11-NEXT:    call void @_ZN1SIiEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[TEST]])
1263 // CHECK11-NEXT:    store i32 0, ptr [[T_VAR]], align 4
1264 // CHECK11-NEXT:    call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[VEC]], ptr align 4 @__const._Z5tmainIiET_v.vec, i32 8, i1 false)
1265 // CHECK11-NEXT:    call void @_ZN1SIiEC1Ei(ptr noundef nonnull align 4 dereferenceable(4) [[S_ARR]], i32 noundef 1)
1266 // CHECK11-NEXT:    [[ARRAYINIT_ELEMENT:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[S_ARR]], i32 1
1267 // CHECK11-NEXT:    call void @_ZN1SIiEC1Ei(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYINIT_ELEMENT]], i32 noundef 2)
1268 // CHECK11-NEXT:    store ptr [[TEST]], ptr [[VAR]], align 4
1269 // CHECK11-NEXT:    store ptr undef, ptr [[_TMP1]], align 4
1270 // CHECK11-NEXT:    [[TMP0:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0
1271 // CHECK11-NEXT:    store i32 3, ptr [[TMP0]], align 4
1272 // CHECK11-NEXT:    [[TMP1:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1
1273 // CHECK11-NEXT:    store i32 0, ptr [[TMP1]], align 4
1274 // CHECK11-NEXT:    [[TMP2:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2
1275 // CHECK11-NEXT:    store ptr null, ptr [[TMP2]], align 4
1276 // CHECK11-NEXT:    [[TMP3:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3
1277 // CHECK11-NEXT:    store ptr null, ptr [[TMP3]], align 4
1278 // CHECK11-NEXT:    [[TMP4:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4
1279 // CHECK11-NEXT:    store ptr null, ptr [[TMP4]], align 4
1280 // CHECK11-NEXT:    [[TMP5:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5
1281 // CHECK11-NEXT:    store ptr null, ptr [[TMP5]], align 4
1282 // CHECK11-NEXT:    [[TMP6:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6
1283 // CHECK11-NEXT:    store ptr null, ptr [[TMP6]], align 4
1284 // CHECK11-NEXT:    [[TMP7:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7
1285 // CHECK11-NEXT:    store ptr null, ptr [[TMP7]], align 4
1286 // CHECK11-NEXT:    [[TMP8:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8
1287 // CHECK11-NEXT:    store i64 2, ptr [[TMP8]], align 8
1288 // CHECK11-NEXT:    [[TMP9:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9
1289 // CHECK11-NEXT:    store i64 0, ptr [[TMP9]], align 8
1290 // CHECK11-NEXT:    [[TMP10:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10
1291 // CHECK11-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP10]], align 4
1292 // CHECK11-NEXT:    [[TMP11:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11
1293 // CHECK11-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP11]], align 4
1294 // CHECK11-NEXT:    [[TMP12:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12
1295 // CHECK11-NEXT:    store i32 0, ptr [[TMP12]], align 4
1296 // CHECK11-NEXT:    [[TMP13:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB2]], i64 -1, i32 0, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49.region_id, ptr [[KERNEL_ARGS]])
1297 // CHECK11-NEXT:    [[TMP14:%.*]] = icmp ne i32 [[TMP13]], 0
1298 // CHECK11-NEXT:    br i1 [[TMP14]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
1299 // CHECK11:       omp_offload.failed:
1300 // CHECK11-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49() #[[ATTR4]]
1301 // CHECK11-NEXT:    br label [[OMP_OFFLOAD_CONT]]
1302 // CHECK11:       omp_offload.cont:
1303 // CHECK11-NEXT:    store i32 0, ptr [[RETVAL]], align 4
1304 // CHECK11-NEXT:    [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i32 0, i32 0
1305 // CHECK11-NEXT:    [[TMP15:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN]], i32 2
1306 // CHECK11-NEXT:    br label [[ARRAYDESTROY_BODY:%.*]]
1307 // CHECK11:       arraydestroy.body:
1308 // CHECK11-NEXT:    [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP15]], [[OMP_OFFLOAD_CONT]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]
1309 // CHECK11-NEXT:    [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i32 -1
1310 // CHECK11-NEXT:    call void @_ZN1SIiED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]
1311 // CHECK11-NEXT:    [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN]]
1312 // CHECK11-NEXT:    br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE2:%.*]], label [[ARRAYDESTROY_BODY]]
1313 // CHECK11:       arraydestroy.done2:
1314 // CHECK11-NEXT:    call void @_ZN1SIiED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[TEST]]) #[[ATTR4]]
1315 // CHECK11-NEXT:    [[TMP16:%.*]] = load i32, ptr [[RETVAL]], align 4
1316 // CHECK11-NEXT:    ret i32 [[TMP16]]
1317 //
1318 //
1319 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC2Ev
1320 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1321 // CHECK11-NEXT:  entry:
1322 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1323 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1324 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1325 // CHECK11-NEXT:    [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0
1326 // CHECK11-NEXT:    store float 0.000000e+00, ptr [[F]], align 4
1327 // CHECK11-NEXT:    ret void
1328 //
1329 //
1330 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfEC2Ef
1331 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]], float noundef [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1332 // CHECK11-NEXT:  entry:
1333 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1334 // CHECK11-NEXT:    [[A_ADDR:%.*]] = alloca float, align 4
1335 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1336 // CHECK11-NEXT:    store float [[A]], ptr [[A_ADDR]], align 4
1337 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1338 // CHECK11-NEXT:    [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0
1339 // CHECK11-NEXT:    [[TMP0:%.*]] = load float, ptr [[A_ADDR]], align 4
1340 // CHECK11-NEXT:    store float [[TMP0]], ptr [[F]], align 4
1341 // CHECK11-NEXT:    ret void
1342 //
1343 //
1344 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIfED2Ev
1345 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1346 // CHECK11-NEXT:  entry:
1347 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1348 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1349 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1350 // CHECK11-NEXT:    ret void
1351 //
1352 //
1353 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC1Ev
1354 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1355 // CHECK11-NEXT:  entry:
1356 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1357 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1358 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1359 // CHECK11-NEXT:    call void @_ZN1SIiEC2Ev(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]])
1360 // CHECK11-NEXT:    ret void
1361 //
1362 //
1363 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC1Ei
1364 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]], i32 noundef [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1365 // CHECK11-NEXT:  entry:
1366 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1367 // CHECK11-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4
1368 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1369 // CHECK11-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 4
1370 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1371 // CHECK11-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4
1372 // CHECK11-NEXT:    call void @_ZN1SIiEC2Ei(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]], i32 noundef [[TMP0]])
1373 // CHECK11-NEXT:    ret void
1374 //
1375 //
1376 // CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49
1377 // CHECK11-SAME: () #[[ATTR3]] {
1378 // CHECK11-NEXT:  entry:
1379 // CHECK11-NEXT:    call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @[[GLOB2]], i32 0, ptr @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49.omp_outlined)
1380 // CHECK11-NEXT:    ret void
1381 //
1382 //
1383 // CHECK11-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z5tmainIiET_v_l49.omp_outlined
1384 // CHECK11-SAME: (ptr noalias noundef [[DOTGLOBAL_TID_:%.*]], ptr noalias noundef [[DOTBOUND_TID_:%.*]]) #[[ATTR3]] {
1385 // CHECK11-NEXT:  entry:
1386 // CHECK11-NEXT:    [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 4
1387 // CHECK11-NEXT:    [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 4
1388 // CHECK11-NEXT:    [[DOTOMP_IV:%.*]] = alloca i32, align 4
1389 // CHECK11-NEXT:    [[TMP:%.*]] = alloca i32, align 4
1390 // CHECK11-NEXT:    [[_TMP1:%.*]] = alloca ptr, align 4
1391 // CHECK11-NEXT:    [[DOTOMP_LB:%.*]] = alloca i32, align 4
1392 // CHECK11-NEXT:    [[DOTOMP_UB:%.*]] = alloca i32, align 4
1393 // CHECK11-NEXT:    [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
1394 // CHECK11-NEXT:    [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
1395 // CHECK11-NEXT:    [[T_VAR:%.*]] = alloca i32, align 4
1396 // CHECK11-NEXT:    [[VEC:%.*]] = alloca [2 x i32], align 4
1397 // CHECK11-NEXT:    [[S_ARR:%.*]] = alloca [2 x %struct.S.0], align 4
1398 // CHECK11-NEXT:    [[VAR:%.*]] = alloca [[STRUCT_S_0:%.*]], align 4
1399 // CHECK11-NEXT:    [[_TMP2:%.*]] = alloca ptr, align 4
1400 // CHECK11-NEXT:    [[I:%.*]] = alloca i32, align 4
1401 // CHECK11-NEXT:    store ptr [[DOTGLOBAL_TID_]], ptr [[DOTGLOBAL_TID__ADDR]], align 4
1402 // CHECK11-NEXT:    store ptr [[DOTBOUND_TID_]], ptr [[DOTBOUND_TID__ADDR]], align 4
1403 // CHECK11-NEXT:    store ptr undef, ptr [[_TMP1]], align 4
1404 // CHECK11-NEXT:    store i32 0, ptr [[DOTOMP_LB]], align 4
1405 // CHECK11-NEXT:    store i32 1, ptr [[DOTOMP_UB]], align 4
1406 // CHECK11-NEXT:    store i32 1, ptr [[DOTOMP_STRIDE]], align 4
1407 // CHECK11-NEXT:    store i32 0, ptr [[DOTOMP_IS_LAST]], align 4
1408 // CHECK11-NEXT:    [[ARRAY_BEGIN:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i32 0, i32 0
1409 // CHECK11-NEXT:    [[ARRAYCTOR_END:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN]], i32 2
1410 // CHECK11-NEXT:    br label [[ARRAYCTOR_LOOP:%.*]]
1411 // CHECK11:       arrayctor.loop:
1412 // CHECK11-NEXT:    [[ARRAYCTOR_CUR:%.*]] = phi ptr [ [[ARRAY_BEGIN]], [[ENTRY:%.*]] ], [ [[ARRAYCTOR_NEXT:%.*]], [[ARRAYCTOR_LOOP]] ]
1413 // CHECK11-NEXT:    call void @_ZN1SIiEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYCTOR_CUR]])
1414 // CHECK11-NEXT:    [[ARRAYCTOR_NEXT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYCTOR_CUR]], i32 1
1415 // CHECK11-NEXT:    [[ARRAYCTOR_DONE:%.*]] = icmp eq ptr [[ARRAYCTOR_NEXT]], [[ARRAYCTOR_END]]
1416 // CHECK11-NEXT:    br i1 [[ARRAYCTOR_DONE]], label [[ARRAYCTOR_CONT:%.*]], label [[ARRAYCTOR_LOOP]]
1417 // CHECK11:       arrayctor.cont:
1418 // CHECK11-NEXT:    call void @_ZN1SIiEC1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[VAR]])
1419 // CHECK11-NEXT:    store ptr [[VAR]], ptr [[_TMP2]], align 4
1420 // CHECK11-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4
1421 // CHECK11-NEXT:    [[TMP1:%.*]] = load i32, ptr [[TMP0]], align 4
1422 // CHECK11-NEXT:    call void @__kmpc_for_static_init_4(ptr @[[GLOB1]], i32 [[TMP1]], i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
1423 // CHECK11-NEXT:    [[TMP2:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
1424 // CHECK11-NEXT:    [[CMP:%.*]] = icmp sgt i32 [[TMP2]], 1
1425 // CHECK11-NEXT:    br i1 [[CMP]], label [[COND_TRUE:%.*]], label [[COND_FALSE:%.*]]
1426 // CHECK11:       cond.true:
1427 // CHECK11-NEXT:    br label [[COND_END:%.*]]
1428 // CHECK11:       cond.false:
1429 // CHECK11-NEXT:    [[TMP3:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
1430 // CHECK11-NEXT:    br label [[COND_END]]
1431 // CHECK11:       cond.end:
1432 // CHECK11-NEXT:    [[COND:%.*]] = phi i32 [ 1, [[COND_TRUE]] ], [ [[TMP3]], [[COND_FALSE]] ]
1433 // CHECK11-NEXT:    store i32 [[COND]], ptr [[DOTOMP_UB]], align 4
1434 // CHECK11-NEXT:    [[TMP4:%.*]] = load i32, ptr [[DOTOMP_LB]], align 4
1435 // CHECK11-NEXT:    store i32 [[TMP4]], ptr [[DOTOMP_IV]], align 4
1436 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_COND:%.*]]
1437 // CHECK11:       omp.inner.for.cond:
1438 // CHECK11-NEXT:    [[TMP5:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
1439 // CHECK11-NEXT:    [[TMP6:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
1440 // CHECK11-NEXT:    [[CMP3:%.*]] = icmp sle i32 [[TMP5]], [[TMP6]]
1441 // CHECK11-NEXT:    br i1 [[CMP3]], label [[OMP_INNER_FOR_BODY:%.*]], label [[OMP_INNER_FOR_COND_CLEANUP:%.*]]
1442 // CHECK11:       omp.inner.for.cond.cleanup:
1443 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_END:%.*]]
1444 // CHECK11:       omp.inner.for.body:
1445 // CHECK11-NEXT:    [[TMP7:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
1446 // CHECK11-NEXT:    [[MUL:%.*]] = mul nsw i32 [[TMP7]], 1
1447 // CHECK11-NEXT:    [[ADD:%.*]] = add nsw i32 0, [[MUL]]
1448 // CHECK11-NEXT:    store i32 [[ADD]], ptr [[I]], align 4
1449 // CHECK11-NEXT:    [[TMP8:%.*]] = load i32, ptr [[T_VAR]], align 4
1450 // CHECK11-NEXT:    [[TMP9:%.*]] = load i32, ptr [[I]], align 4
1451 // CHECK11-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x i32], ptr [[VEC]], i32 0, i32 [[TMP9]]
1452 // CHECK11-NEXT:    store i32 [[TMP8]], ptr [[ARRAYIDX]], align 4
1453 // CHECK11-NEXT:    [[TMP10:%.*]] = load ptr, ptr [[_TMP2]], align 4
1454 // CHECK11-NEXT:    [[TMP11:%.*]] = load i32, ptr [[I]], align 4
1455 // CHECK11-NEXT:    [[ARRAYIDX4:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i32 0, i32 [[TMP11]]
1456 // CHECK11-NEXT:    call void @llvm.memcpy.p0.p0.i32(ptr align 4 [[ARRAYIDX4]], ptr align 4 [[TMP10]], i32 4, i1 false)
1457 // CHECK11-NEXT:    br label [[OMP_BODY_CONTINUE:%.*]]
1458 // CHECK11:       omp.body.continue:
1459 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_INC:%.*]]
1460 // CHECK11:       omp.inner.for.inc:
1461 // CHECK11-NEXT:    [[TMP12:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
1462 // CHECK11-NEXT:    [[ADD5:%.*]] = add nsw i32 [[TMP12]], 1
1463 // CHECK11-NEXT:    store i32 [[ADD5]], ptr [[DOTOMP_IV]], align 4
1464 // CHECK11-NEXT:    br label [[OMP_INNER_FOR_COND]]
1465 // CHECK11:       omp.inner.for.end:
1466 // CHECK11-NEXT:    br label [[OMP_LOOP_EXIT:%.*]]
1467 // CHECK11:       omp.loop.exit:
1468 // CHECK11-NEXT:    [[TMP13:%.*]] = load ptr, ptr [[DOTGLOBAL_TID__ADDR]], align 4
1469 // CHECK11-NEXT:    [[TMP14:%.*]] = load i32, ptr [[TMP13]], align 4
1470 // CHECK11-NEXT:    call void @__kmpc_for_static_fini(ptr @[[GLOB1]], i32 [[TMP14]])
1471 // CHECK11-NEXT:    call void @_ZN1SIiED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[VAR]]) #[[ATTR4]]
1472 // CHECK11-NEXT:    [[ARRAY_BEGIN6:%.*]] = getelementptr inbounds [2 x %struct.S.0], ptr [[S_ARR]], i32 0, i32 0
1473 // CHECK11-NEXT:    [[TMP15:%.*]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAY_BEGIN6]], i32 2
1474 // CHECK11-NEXT:    br label [[ARRAYDESTROY_BODY:%.*]]
1475 // CHECK11:       arraydestroy.body:
1476 // CHECK11-NEXT:    [[ARRAYDESTROY_ELEMENTPAST:%.*]] = phi ptr [ [[TMP15]], [[OMP_LOOP_EXIT]] ], [ [[ARRAYDESTROY_ELEMENT:%.*]], [[ARRAYDESTROY_BODY]] ]
1477 // CHECK11-NEXT:    [[ARRAYDESTROY_ELEMENT]] = getelementptr inbounds [[STRUCT_S_0]], ptr [[ARRAYDESTROY_ELEMENTPAST]], i32 -1
1478 // CHECK11-NEXT:    call void @_ZN1SIiED1Ev(ptr noundef nonnull align 4 dereferenceable(4) [[ARRAYDESTROY_ELEMENT]]) #[[ATTR4]]
1479 // CHECK11-NEXT:    [[ARRAYDESTROY_DONE:%.*]] = icmp eq ptr [[ARRAYDESTROY_ELEMENT]], [[ARRAY_BEGIN6]]
1480 // CHECK11-NEXT:    br i1 [[ARRAYDESTROY_DONE]], label [[ARRAYDESTROY_DONE7:%.*]], label [[ARRAYDESTROY_BODY]]
1481 // CHECK11:       arraydestroy.done7:
1482 // CHECK11-NEXT:    ret void
1483 //
1484 //
1485 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiED1Ev
1486 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1487 // CHECK11-NEXT:  entry:
1488 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1489 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1490 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1491 // CHECK11-NEXT:    call void @_ZN1SIiED2Ev(ptr noundef nonnull align 4 dereferenceable(4) [[THIS1]]) #[[ATTR4]]
1492 // CHECK11-NEXT:    ret void
1493 //
1494 //
1495 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC2Ev
1496 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1497 // CHECK11-NEXT:  entry:
1498 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1499 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1500 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1501 // CHECK11-NEXT:    [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 0
1502 // CHECK11-NEXT:    store i32 0, ptr [[F]], align 4
1503 // CHECK11-NEXT:    ret void
1504 //
1505 //
1506 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiEC2Ei
1507 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]], i32 noundef [[A:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1508 // CHECK11-NEXT:  entry:
1509 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1510 // CHECK11-NEXT:    [[A_ADDR:%.*]] = alloca i32, align 4
1511 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1512 // CHECK11-NEXT:    store i32 [[A]], ptr [[A_ADDR]], align 4
1513 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1514 // CHECK11-NEXT:    [[F:%.*]] = getelementptr inbounds nuw [[STRUCT_S_0:%.*]], ptr [[THIS1]], i32 0, i32 0
1515 // CHECK11-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A_ADDR]], align 4
1516 // CHECK11-NEXT:    store i32 [[TMP0]], ptr [[F]], align 4
1517 // CHECK11-NEXT:    ret void
1518 //
1519 //
1520 // CHECK11-LABEL: define {{[^@]+}}@_ZN1SIiED2Ev
1521 // CHECK11-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[THIS:%.*]]) unnamed_addr #[[ATTR1]] comdat align 2 {
1522 // CHECK11-NEXT:  entry:
1523 // CHECK11-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 4
1524 // CHECK11-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 4
1525 // CHECK11-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 4
1526 // CHECK11-NEXT:    ret void
1527 //