| // RUN: hlo_to_llvm_ir %s | FileCheck %s |
| |
| // NOTE: Assertions have been autogenerated by utils/generate-test-checks.py |
| |
| // CHECK-LABEL: entry: |
| // CHECK: %[[VAL_0:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_1:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_2:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_3:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_4:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_5:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_6:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_7:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_8:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_9:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_10:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_11:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_12:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_13:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_14:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_15:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_16:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_17:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_18:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_19:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_20:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_21:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_22:.*]] = alloca i32, align 4 |
| // CHECK: %[[VAL_23:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_24:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_25:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_26:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_27:.*]] = getelementptr inbounds i8, i8* %[[VAL_28:.*]], i64 0 |
| // CHECK: %[[VAL_29:.*]] = bitcast i8* %[[VAL_27]] to [2 x [32 x [32 x float]]]* |
| // CHECK: %[[VAL_30:.*]] = getelementptr inbounds i8, i8* %[[VAL_31:.*]], i64 0 |
| // CHECK: %[[VAL_32:.*]] = bitcast i8* %[[VAL_30]] to float* |
| // CHECK: %[[VAL_33:.*]] = getelementptr inbounds i8, i8* %[[VAL_34:.*]], i64 0 |
| // CHECK: %[[VAL_35:.*]] = bitcast i8* %[[VAL_33]] to float* |
| // CHECK: %[[VAL_36:.*]] = getelementptr inbounds i8, i8* %[[VAL_37:.*]], i64 0 |
| // CHECK: %[[VAL_38:.*]] = bitcast i8* %[[VAL_36]] to [2 x [32 x float]]* |
| // CHECK: %[[VAL_39:.*]] = getelementptr inbounds i8, i8* %[[VAL_40:.*]], i64 0 |
| // CHECK: %[[VAL_41:.*]] = bitcast i8* %[[VAL_39]] to [2 x [32 x float]]* |
| // CHECK: %[[VAL_42:.*]] = call i32 @llvm.nvvm.read.ptx.sreg.ctaid.y(), !range !2 |
| // CHECK: %[[VAL_43:.*]] = icmp eq i32 %[[VAL_42]], 0 |
| // CHECK: br i1 %[[VAL_43]], label %[[VAL_44:.*]], label %[[VAL_45:.*]] |
| // CHECK: reduce-group-0-after: ; preds = %[[VAL_46:.*]], %[[VAL_47:.*]] |
| // CHECK: ret void |
| // CHECK: reduce-group-0-true: ; preds = %[[VAL_47]] |
| // CHECK: %[[VAL_48:.*]] = load float, float* %[[VAL_32]], align 4, !invariant.load !3 |
| // CHECK: %[[VAL_49:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: store float %[[VAL_48]], float* %[[VAL_49]], align 4 |
| // CHECK: %[[VAL_50:.*]] = load float, float* %[[VAL_35]], align 4, !invariant.load !3 |
| // CHECK: %[[VAL_51:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: store float %[[VAL_50]], float* %[[VAL_51]], align 4 |
| // CHECK: %[[VAL_52:.*]] = call i32 @llvm.nvvm.read.ptx.sreg.tid.x(), !range !4 |
| // CHECK: %[[VAL_53:.*]] = urem i32 %[[VAL_52]], 32 |
| // CHECK: %[[VAL_54:.*]] = udiv i32 %[[VAL_52]], 32 |
| // CHECK: %[[VAL_55:.*]] = urem i32 %[[VAL_52]], 32 |
| // CHECK: %[[VAL_56:.*]] = call i32 @llvm.nvvm.read.ptx.sreg.ctaid.x(), !range !5 |
| // CHECK: %[[VAL_57:.*]] = udiv i32 %[[VAL_56]], 1 |
| // CHECK: %[[VAL_58:.*]] = urem i32 %[[VAL_57]], 1 |
| // CHECK: %[[VAL_59:.*]] = udiv i32 %[[VAL_56]], 1 |
| // CHECK: %[[VAL_60:.*]] = urem i32 %[[VAL_59]], 64 |
| // CHECK: %[[VAL_61:.*]] = udiv i32 %[[VAL_56]], 64 |
| // CHECK: %[[VAL_62:.*]] = mul i32 %[[VAL_61]], 1 |
| // CHECK: %[[VAL_63:.*]] = icmp eq i32 %[[VAL_60]], 63 |
| // CHECK: %[[VAL_64:.*]] = select i1 %[[VAL_63]], i32 1, i32 1 |
| // CHECK: %[[VAL_65:.*]] = icmp eq i32 %[[VAL_58]], 0 |
| // CHECK: %[[VAL_66:.*]] = select i1 %[[VAL_65]], i32 32, i32 2048 |
| // CHECK: %[[VAL_67:.*]] = mul i32 %[[VAL_60]], 1 |
| // CHECK: %[[VAL_68:.*]] = mul i32 %[[VAL_58]], 2048 |
| // CHECK: %[[VAL_69:.*]] = mul i32 %[[VAL_53]], 2 |
| // CHECK: %[[VAL_70:.*]] = add i32 %[[VAL_68]], %[[VAL_69]] |
| // CHECK: store i32 %[[VAL_54]], i32* %[[VAL_22]], align 4 |
| // CHECK: br label %[[VAL_71:.*]] |
| // CHECK: output_y_in_tile.loop_header: ; preds = %[[VAL_72:.*]], %[[VAL_44]] |
| // CHECK: %[[VAL_73:.*]] = load i32, i32* %[[VAL_22]], align 4 |
| // CHECK: %[[VAL_74:.*]] = icmp uge i32 %[[VAL_73]], %[[VAL_64]] |
| // CHECK: br i1 %[[VAL_74]], label %[[VAL_75:.*]], label %[[VAL_76:.*]] |
| // CHECK: output_y_in_tile.loop_body: ; preds = %[[VAL_71]] |
| // CHECK: %[[VAL_77:.*]] = add nuw nsw i32 %[[VAL_73]], 1 |
| // CHECK: store i32 %[[VAL_77]], i32* %[[VAL_22]], align 4 |
| // CHECK: %[[VAL_78:.*]] = icmp eq i32 %[[VAL_73]], %[[VAL_54]] |
| // CHECK: %[[VAL_79:.*]] = icmp eq i32 2048, %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_79]], label %[[VAL_80:.*]], label %[[VAL_81:.*]] |
| // CHECK: output_is_full_tile-after: ; preds = %[[VAL_82:.*]], %[[VAL_80]] |
| // CHECK: br label %[[VAL_71]], !llvm.loop !6 |
| // CHECK: output_y_in_tile.loop_exit: ; preds = %[[VAL_71]] |
| // CHECK: %[[VAL_83:.*]] = call i32 @llvm.nvvm.read.ptx.sreg.tid.x(), !range !4 |
| // CHECK: %[[VAL_84:.*]] = urem i32 %[[VAL_83]], 32 |
| // CHECK: %[[VAL_85:.*]] = udiv i32 %[[VAL_83]], 32 |
| // CHECK: %[[VAL_86:.*]] = urem i32 %[[VAL_83]], 32 |
| // CHECK: %[[VAL_87:.*]] = mul i32 %[[VAL_84]], 2 |
| // CHECK: %[[VAL_88:.*]] = add i32 %[[VAL_67]], %[[VAL_85]] |
| // CHECK: %[[VAL_89:.*]] = add i32 %[[VAL_68]], %[[VAL_87]] |
| // CHECK: %[[VAL_90:.*]] = add i32 %[[VAL_89]], 0 |
| // CHECK: %[[VAL_91:.*]] = udiv i32 %[[VAL_88]], 1 |
| // CHECK: %[[VAL_92:.*]] = urem i32 %[[VAL_91]], 32 |
| // CHECK: %[[VAL_93:.*]] = udiv i32 %[[VAL_88]], 32 |
| // CHECK: %[[VAL_94:.*]] = getelementptr inbounds [2 x [32 x float]], [2 x [32 x float]]* %[[VAL_38]], i32 0, i32 %[[VAL_93]], i32 %[[VAL_92]] |
| // CHECK: %[[VAL_95:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: %[[VAL_96:.*]] = load float, float* %[[VAL_95]], align 4 |
| // CHECK: %[[VAL_97:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_96]], i32 16, i32 31) |
| // CHECK: store float %[[VAL_97]], float* %[[VAL_21]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_95]], float* %[[VAL_21]], float* %[[VAL_95]]) |
| // CHECK: %[[VAL_98:.*]] = load float, float* %[[VAL_95]], align 4 |
| // CHECK: %[[VAL_99:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_98]], i32 8, i32 31) |
| // CHECK: store float %[[VAL_99]], float* %[[VAL_20]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_95]], float* %[[VAL_20]], float* %[[VAL_95]]) |
| // CHECK: %[[VAL_100:.*]] = load float, float* %[[VAL_95]], align 4 |
| // CHECK: %[[VAL_101:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_100]], i32 4, i32 31) |
| // CHECK: store float %[[VAL_101]], float* %[[VAL_19]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_95]], float* %[[VAL_19]], float* %[[VAL_95]]) |
| // CHECK: %[[VAL_102:.*]] = load float, float* %[[VAL_95]], align 4 |
| // CHECK: %[[VAL_103:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_102]], i32 2, i32 31) |
| // CHECK: store float %[[VAL_103]], float* %[[VAL_18]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_95]], float* %[[VAL_18]], float* %[[VAL_95]]) |
| // CHECK: %[[VAL_104:.*]] = load float, float* %[[VAL_95]], align 4 |
| // CHECK: %[[VAL_105:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_104]], i32 1, i32 31) |
| // CHECK: store float %[[VAL_105]], float* %[[VAL_17]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_95]], float* %[[VAL_17]], float* %[[VAL_95]]) |
| // CHECK: %[[VAL_106:.*]] = udiv i32 %[[VAL_84]], 32 |
| // CHECK: %[[VAL_107:.*]] = icmp eq i32 %[[VAL_86]], 0 |
| // CHECK: br i1 %[[VAL_107]], label %[[VAL_108:.*]], label %[[VAL_109:.*]] |
| // CHECK: intra_warp_reduce_write-after: ; preds = %[[VAL_108]], %[[VAL_75]] |
| // CHECK: call void @llvm.nvvm.barrier0() |
| // CHECK: %[[VAL_110:.*]] = icmp eq i32 %[[VAL_106]], 0 |
| // CHECK: br i1 %[[VAL_110]], label %[[VAL_111:.*]], label %[[VAL_112:.*]] |
| // CHECK: inter_warp_reduce-after: ; preds = %[[VAL_113:.*]], %[[VAL_109]] |
| // CHECK: %[[VAL_114:.*]] = add i32 %[[VAL_89]], 0 |
| // CHECK: %[[VAL_115:.*]] = udiv i32 %[[VAL_88]], 1 |
| // CHECK: %[[VAL_116:.*]] = urem i32 %[[VAL_115]], 32 |
| // CHECK: %[[VAL_117:.*]] = udiv i32 %[[VAL_88]], 32 |
| // CHECK: %[[VAL_118:.*]] = getelementptr inbounds [2 x [32 x float]], [2 x [32 x float]]* %[[VAL_41]], i32 0, i32 %[[VAL_117]], i32 %[[VAL_116]] |
| // CHECK: %[[VAL_119:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: %[[VAL_120:.*]] = load float, float* %[[VAL_119]], align 4 |
| // CHECK: %[[VAL_121:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_120]], i32 16, i32 31) |
| // CHECK: store float %[[VAL_121]], float* %[[VAL_10]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_119]], float* %[[VAL_10]], float* %[[VAL_119]]) |
| // CHECK: %[[VAL_122:.*]] = load float, float* %[[VAL_119]], align 4 |
| // CHECK: %[[VAL_123:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_122]], i32 8, i32 31) |
| // CHECK: store float %[[VAL_123]], float* %[[VAL_9]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_119]], float* %[[VAL_9]], float* %[[VAL_119]]) |
| // CHECK: %[[VAL_124:.*]] = load float, float* %[[VAL_119]], align 4 |
| // CHECK: %[[VAL_125:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_124]], i32 4, i32 31) |
| // CHECK: store float %[[VAL_125]], float* %[[VAL_8]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_119]], float* %[[VAL_8]], float* %[[VAL_119]]) |
| // CHECK: %[[VAL_126:.*]] = load float, float* %[[VAL_119]], align 4 |
| // CHECK: %[[VAL_127:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_126]], i32 2, i32 31) |
| // CHECK: store float %[[VAL_127]], float* %[[VAL_7]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_119]], float* %[[VAL_7]], float* %[[VAL_119]]) |
| // CHECK: %[[VAL_128:.*]] = load float, float* %[[VAL_119]], align 4 |
| // CHECK: %[[VAL_129:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_128]], i32 1, i32 31) |
| // CHECK: store float %[[VAL_129]], float* %[[VAL_6]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_119]], float* %[[VAL_6]], float* %[[VAL_119]]) |
| // CHECK: %[[VAL_130:.*]] = udiv i32 %[[VAL_84]], 32 |
| // CHECK: %[[VAL_131:.*]] = icmp eq i32 %[[VAL_86]], 0 |
| // CHECK: br i1 %[[VAL_131]], label %[[VAL_132:.*]], label %[[VAL_133:.*]] |
| // CHECK: intra_warp_reduce_write-after801: ; preds = %[[VAL_132]], %[[VAL_112]] |
| // CHECK: call void @llvm.nvvm.barrier0() |
| // CHECK: %[[VAL_134:.*]] = icmp eq i32 %[[VAL_130]], 0 |
| // CHECK: br i1 %[[VAL_134]], label %[[VAL_135:.*]], label %[[VAL_46]] |
| // CHECK: inter_warp_reduce-after803: ; preds = %[[VAL_136:.*]], %[[VAL_133]] |
| // CHECK: br label %[[VAL_45]] |
| // CHECK: output_is_full_tile-true: ; preds = %[[VAL_76]] |
| // CHECK: %[[VAL_137:.*]] = add i32 %[[VAL_67]], %[[VAL_73]] |
| // CHECK: %[[VAL_138:.*]] = add i32 0, %[[VAL_69]] |
| // CHECK: %[[VAL_139:.*]] = add i32 %[[VAL_70]], 0 |
| // CHECK: %[[VAL_140:.*]] = mul nuw nsw i32 %[[VAL_139]], 1 |
| // CHECK: %[[VAL_141:.*]] = add nuw nsw i32 0, %[[VAL_140]] |
| // CHECK: %[[VAL_142:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_143:.*]] = add nuw nsw i32 %[[VAL_141]], %[[VAL_142]] |
| // CHECK: %[[VAL_144:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_145:.*]] = add nuw nsw i32 %[[VAL_143]], %[[VAL_144]] |
| // CHECK: %[[VAL_146:.*]] = udiv i32 %[[VAL_145]], 1 |
| // CHECK: %[[VAL_147:.*]] = urem i32 %[[VAL_146]], 32 |
| // CHECK: %[[VAL_148:.*]] = udiv i32 %[[VAL_145]], 32 |
| // CHECK: %[[VAL_149:.*]] = urem i32 %[[VAL_148]], 32 |
| // CHECK: %[[VAL_150:.*]] = udiv i32 %[[VAL_145]], 1024 |
| // CHECK: %[[VAL_151:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_152:.*]] = getelementptr inbounds float, float* %[[VAL_151]], i32 %[[VAL_145]] |
| // CHECK: %[[VAL_153:.*]] = load float, float* %[[VAL_152]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_153]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_154:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_154]], float* %[[VAL_26]], float* %[[VAL_154]]) |
| // CHECK: %[[VAL_155:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_156:.*]] = getelementptr inbounds float, float* %[[VAL_155]], i32 %[[VAL_145]] |
| // CHECK: %[[VAL_157:.*]] = load float, float* %[[VAL_156]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_157]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_158:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_158]], float* %[[VAL_24]], float* %[[VAL_158]]) |
| // CHECK: %[[VAL_159:.*]] = add i32 1, %[[VAL_69]] |
| // CHECK: %[[VAL_160:.*]] = add i32 %[[VAL_70]], 1 |
| // CHECK: %[[VAL_161:.*]] = mul nuw nsw i32 %[[VAL_160]], 1 |
| // CHECK: %[[VAL_162:.*]] = add nuw nsw i32 0, %[[VAL_161]] |
| // CHECK: %[[VAL_163:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_164:.*]] = add nuw nsw i32 %[[VAL_162]], %[[VAL_163]] |
| // CHECK: %[[VAL_165:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_166:.*]] = add nuw nsw i32 %[[VAL_164]], %[[VAL_165]] |
| // CHECK: %[[VAL_167:.*]] = udiv i32 %[[VAL_166]], 1 |
| // CHECK: %[[VAL_168:.*]] = urem i32 %[[VAL_167]], 32 |
| // CHECK: %[[VAL_169:.*]] = udiv i32 %[[VAL_166]], 32 |
| // CHECK: %[[VAL_170:.*]] = urem i32 %[[VAL_169]], 32 |
| // CHECK: %[[VAL_171:.*]] = udiv i32 %[[VAL_166]], 1024 |
| // CHECK: %[[VAL_172:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_173:.*]] = getelementptr inbounds float, float* %[[VAL_172]], i32 %[[VAL_166]] |
| // CHECK: %[[VAL_174:.*]] = load float, float* %[[VAL_173]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_174]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_175:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_175]], float* %[[VAL_26]], float* %[[VAL_175]]) |
| // CHECK: %[[VAL_176:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_177:.*]] = getelementptr inbounds float, float* %[[VAL_176]], i32 %[[VAL_166]] |
| // CHECK: %[[VAL_178:.*]] = load float, float* %[[VAL_177]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_178]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_179:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_179]], float* %[[VAL_24]], float* %[[VAL_179]]) |
| // CHECK: %[[VAL_180:.*]] = add i32 64, %[[VAL_69]] |
| // CHECK: %[[VAL_181:.*]] = add i32 %[[VAL_70]], 64 |
| // CHECK: %[[VAL_182:.*]] = mul nuw nsw i32 %[[VAL_181]], 1 |
| // CHECK: %[[VAL_183:.*]] = add nuw nsw i32 0, %[[VAL_182]] |
| // CHECK: %[[VAL_184:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_185:.*]] = add nuw nsw i32 %[[VAL_183]], %[[VAL_184]] |
| // CHECK: %[[VAL_186:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_187:.*]] = add nuw nsw i32 %[[VAL_185]], %[[VAL_186]] |
| // CHECK: %[[VAL_188:.*]] = udiv i32 %[[VAL_187]], 1 |
| // CHECK: %[[VAL_189:.*]] = urem i32 %[[VAL_188]], 32 |
| // CHECK: %[[VAL_190:.*]] = udiv i32 %[[VAL_187]], 32 |
| // CHECK: %[[VAL_191:.*]] = urem i32 %[[VAL_190]], 32 |
| // CHECK: %[[VAL_192:.*]] = udiv i32 %[[VAL_187]], 1024 |
| // CHECK: %[[VAL_193:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_194:.*]] = getelementptr inbounds float, float* %[[VAL_193]], i32 %[[VAL_187]] |
| // CHECK: %[[VAL_195:.*]] = load float, float* %[[VAL_194]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_195]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_196:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_196]], float* %[[VAL_26]], float* %[[VAL_196]]) |
| // CHECK: %[[VAL_197:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_198:.*]] = getelementptr inbounds float, float* %[[VAL_197]], i32 %[[VAL_187]] |
| // CHECK: %[[VAL_199:.*]] = load float, float* %[[VAL_198]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_199]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_200:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_200]], float* %[[VAL_24]], float* %[[VAL_200]]) |
| // CHECK: %[[VAL_201:.*]] = add i32 65, %[[VAL_69]] |
| // CHECK: %[[VAL_202:.*]] = add i32 %[[VAL_70]], 65 |
| // CHECK: %[[VAL_203:.*]] = mul nuw nsw i32 %[[VAL_202]], 1 |
| // CHECK: %[[VAL_204:.*]] = add nuw nsw i32 0, %[[VAL_203]] |
| // CHECK: %[[VAL_205:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_206:.*]] = add nuw nsw i32 %[[VAL_204]], %[[VAL_205]] |
| // CHECK: %[[VAL_207:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_208:.*]] = add nuw nsw i32 %[[VAL_206]], %[[VAL_207]] |
| // CHECK: %[[VAL_209:.*]] = udiv i32 %[[VAL_208]], 1 |
| // CHECK: %[[VAL_210:.*]] = urem i32 %[[VAL_209]], 32 |
| // CHECK: %[[VAL_211:.*]] = udiv i32 %[[VAL_208]], 32 |
| // CHECK: %[[VAL_212:.*]] = urem i32 %[[VAL_211]], 32 |
| // CHECK: %[[VAL_213:.*]] = udiv i32 %[[VAL_208]], 1024 |
| // CHECK: %[[VAL_214:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_215:.*]] = getelementptr inbounds float, float* %[[VAL_214]], i32 %[[VAL_208]] |
| // CHECK: %[[VAL_216:.*]] = load float, float* %[[VAL_215]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_216]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_217:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_217]], float* %[[VAL_26]], float* %[[VAL_217]]) |
| // CHECK: %[[VAL_218:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_219:.*]] = getelementptr inbounds float, float* %[[VAL_218]], i32 %[[VAL_208]] |
| // CHECK: %[[VAL_220:.*]] = load float, float* %[[VAL_219]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_220]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_221:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_221]], float* %[[VAL_24]], float* %[[VAL_221]]) |
| // CHECK: %[[VAL_222:.*]] = add i32 128, %[[VAL_69]] |
| // CHECK: %[[VAL_223:.*]] = add i32 %[[VAL_70]], 128 |
| // CHECK: %[[VAL_224:.*]] = mul nuw nsw i32 %[[VAL_223]], 1 |
| // CHECK: %[[VAL_225:.*]] = add nuw nsw i32 0, %[[VAL_224]] |
| // CHECK: %[[VAL_226:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_227:.*]] = add nuw nsw i32 %[[VAL_225]], %[[VAL_226]] |
| // CHECK: %[[VAL_228:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_229:.*]] = add nuw nsw i32 %[[VAL_227]], %[[VAL_228]] |
| // CHECK: %[[VAL_230:.*]] = udiv i32 %[[VAL_229]], 1 |
| // CHECK: %[[VAL_231:.*]] = urem i32 %[[VAL_230]], 32 |
| // CHECK: %[[VAL_232:.*]] = udiv i32 %[[VAL_229]], 32 |
| // CHECK: %[[VAL_233:.*]] = urem i32 %[[VAL_232]], 32 |
| // CHECK: %[[VAL_234:.*]] = udiv i32 %[[VAL_229]], 1024 |
| // CHECK: %[[VAL_235:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_236:.*]] = getelementptr inbounds float, float* %[[VAL_235]], i32 %[[VAL_229]] |
| // CHECK: %[[VAL_237:.*]] = load float, float* %[[VAL_236]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_237]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_238:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_238]], float* %[[VAL_26]], float* %[[VAL_238]]) |
| // CHECK: %[[VAL_239:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_240:.*]] = getelementptr inbounds float, float* %[[VAL_239]], i32 %[[VAL_229]] |
| // CHECK: %[[VAL_241:.*]] = load float, float* %[[VAL_240]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_241]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_242:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_242]], float* %[[VAL_24]], float* %[[VAL_242]]) |
| // CHECK: %[[VAL_243:.*]] = add i32 129, %[[VAL_69]] |
| // CHECK: %[[VAL_244:.*]] = add i32 %[[VAL_70]], 129 |
| // CHECK: %[[VAL_245:.*]] = mul nuw nsw i32 %[[VAL_244]], 1 |
| // CHECK: %[[VAL_246:.*]] = add nuw nsw i32 0, %[[VAL_245]] |
| // CHECK: %[[VAL_247:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_248:.*]] = add nuw nsw i32 %[[VAL_246]], %[[VAL_247]] |
| // CHECK: %[[VAL_249:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_250:.*]] = add nuw nsw i32 %[[VAL_248]], %[[VAL_249]] |
| // CHECK: %[[VAL_251:.*]] = udiv i32 %[[VAL_250]], 1 |
| // CHECK: %[[VAL_252:.*]] = urem i32 %[[VAL_251]], 32 |
| // CHECK: %[[VAL_253:.*]] = udiv i32 %[[VAL_250]], 32 |
| // CHECK: %[[VAL_254:.*]] = urem i32 %[[VAL_253]], 32 |
| // CHECK: %[[VAL_255:.*]] = udiv i32 %[[VAL_250]], 1024 |
| // CHECK: %[[VAL_256:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_257:.*]] = getelementptr inbounds float, float* %[[VAL_256]], i32 %[[VAL_250]] |
| // CHECK: %[[VAL_258:.*]] = load float, float* %[[VAL_257]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_258]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_259:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_259]], float* %[[VAL_26]], float* %[[VAL_259]]) |
| // CHECK: %[[VAL_260:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_261:.*]] = getelementptr inbounds float, float* %[[VAL_260]], i32 %[[VAL_250]] |
| // CHECK: %[[VAL_262:.*]] = load float, float* %[[VAL_261]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_262]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_263:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_263]], float* %[[VAL_24]], float* %[[VAL_263]]) |
| // CHECK: %[[VAL_264:.*]] = add i32 192, %[[VAL_69]] |
| // CHECK: %[[VAL_265:.*]] = add i32 %[[VAL_70]], 192 |
| // CHECK: %[[VAL_266:.*]] = mul nuw nsw i32 %[[VAL_265]], 1 |
| // CHECK: %[[VAL_267:.*]] = add nuw nsw i32 0, %[[VAL_266]] |
| // CHECK: %[[VAL_268:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_269:.*]] = add nuw nsw i32 %[[VAL_267]], %[[VAL_268]] |
| // CHECK: %[[VAL_270:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_271:.*]] = add nuw nsw i32 %[[VAL_269]], %[[VAL_270]] |
| // CHECK: %[[VAL_272:.*]] = udiv i32 %[[VAL_271]], 1 |
| // CHECK: %[[VAL_273:.*]] = urem i32 %[[VAL_272]], 32 |
| // CHECK: %[[VAL_274:.*]] = udiv i32 %[[VAL_271]], 32 |
| // CHECK: %[[VAL_275:.*]] = urem i32 %[[VAL_274]], 32 |
| // CHECK: %[[VAL_276:.*]] = udiv i32 %[[VAL_271]], 1024 |
| // CHECK: %[[VAL_277:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_278:.*]] = getelementptr inbounds float, float* %[[VAL_277]], i32 %[[VAL_271]] |
| // CHECK: %[[VAL_279:.*]] = load float, float* %[[VAL_278]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_279]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_280:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_280]], float* %[[VAL_26]], float* %[[VAL_280]]) |
| // CHECK: %[[VAL_281:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_282:.*]] = getelementptr inbounds float, float* %[[VAL_281]], i32 %[[VAL_271]] |
| // CHECK: %[[VAL_283:.*]] = load float, float* %[[VAL_282]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_283]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_284:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_284]], float* %[[VAL_24]], float* %[[VAL_284]]) |
| // CHECK: %[[VAL_285:.*]] = add i32 193, %[[VAL_69]] |
| // CHECK: %[[VAL_286:.*]] = add i32 %[[VAL_70]], 193 |
| // CHECK: %[[VAL_287:.*]] = mul nuw nsw i32 %[[VAL_286]], 1 |
| // CHECK: %[[VAL_288:.*]] = add nuw nsw i32 0, %[[VAL_287]] |
| // CHECK: %[[VAL_289:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_290:.*]] = add nuw nsw i32 %[[VAL_288]], %[[VAL_289]] |
| // CHECK: %[[VAL_291:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_292:.*]] = add nuw nsw i32 %[[VAL_290]], %[[VAL_291]] |
| // CHECK: %[[VAL_293:.*]] = udiv i32 %[[VAL_292]], 1 |
| // CHECK: %[[VAL_294:.*]] = urem i32 %[[VAL_293]], 32 |
| // CHECK: %[[VAL_295:.*]] = udiv i32 %[[VAL_292]], 32 |
| // CHECK: %[[VAL_296:.*]] = urem i32 %[[VAL_295]], 32 |
| // CHECK: %[[VAL_297:.*]] = udiv i32 %[[VAL_292]], 1024 |
| // CHECK: %[[VAL_298:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_299:.*]] = getelementptr inbounds float, float* %[[VAL_298]], i32 %[[VAL_292]] |
| // CHECK: %[[VAL_300:.*]] = load float, float* %[[VAL_299]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_300]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_301:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_301]], float* %[[VAL_26]], float* %[[VAL_301]]) |
| // CHECK: %[[VAL_302:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_303:.*]] = getelementptr inbounds float, float* %[[VAL_302]], i32 %[[VAL_292]] |
| // CHECK: %[[VAL_304:.*]] = load float, float* %[[VAL_303]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_304]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_305:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_305]], float* %[[VAL_24]], float* %[[VAL_305]]) |
| // CHECK: %[[VAL_306:.*]] = add i32 256, %[[VAL_69]] |
| // CHECK: %[[VAL_307:.*]] = add i32 %[[VAL_70]], 256 |
| // CHECK: %[[VAL_308:.*]] = mul nuw nsw i32 %[[VAL_307]], 1 |
| // CHECK: %[[VAL_309:.*]] = add nuw nsw i32 0, %[[VAL_308]] |
| // CHECK: %[[VAL_310:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_311:.*]] = add nuw nsw i32 %[[VAL_309]], %[[VAL_310]] |
| // CHECK: %[[VAL_312:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_313:.*]] = add nuw nsw i32 %[[VAL_311]], %[[VAL_312]] |
| // CHECK: %[[VAL_314:.*]] = udiv i32 %[[VAL_313]], 1 |
| // CHECK: %[[VAL_315:.*]] = urem i32 %[[VAL_314]], 32 |
| // CHECK: %[[VAL_316:.*]] = udiv i32 %[[VAL_313]], 32 |
| // CHECK: %[[VAL_317:.*]] = urem i32 %[[VAL_316]], 32 |
| // CHECK: %[[VAL_318:.*]] = udiv i32 %[[VAL_313]], 1024 |
| // CHECK: %[[VAL_319:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_320:.*]] = getelementptr inbounds float, float* %[[VAL_319]], i32 %[[VAL_313]] |
| // CHECK: %[[VAL_321:.*]] = load float, float* %[[VAL_320]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_321]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_322:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_322]], float* %[[VAL_26]], float* %[[VAL_322]]) |
| // CHECK: %[[VAL_323:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_324:.*]] = getelementptr inbounds float, float* %[[VAL_323]], i32 %[[VAL_313]] |
| // CHECK: %[[VAL_325:.*]] = load float, float* %[[VAL_324]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_325]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_326:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_326]], float* %[[VAL_24]], float* %[[VAL_326]]) |
| // CHECK: %[[VAL_327:.*]] = add i32 257, %[[VAL_69]] |
| // CHECK: %[[VAL_328:.*]] = add i32 %[[VAL_70]], 257 |
| // CHECK: %[[VAL_329:.*]] = mul nuw nsw i32 %[[VAL_328]], 1 |
| // CHECK: %[[VAL_330:.*]] = add nuw nsw i32 0, %[[VAL_329]] |
| // CHECK: %[[VAL_331:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_332:.*]] = add nuw nsw i32 %[[VAL_330]], %[[VAL_331]] |
| // CHECK: %[[VAL_333:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_334:.*]] = add nuw nsw i32 %[[VAL_332]], %[[VAL_333]] |
| // CHECK: %[[VAL_335:.*]] = udiv i32 %[[VAL_334]], 1 |
| // CHECK: %[[VAL_336:.*]] = urem i32 %[[VAL_335]], 32 |
| // CHECK: %[[VAL_337:.*]] = udiv i32 %[[VAL_334]], 32 |
| // CHECK: %[[VAL_338:.*]] = urem i32 %[[VAL_337]], 32 |
| // CHECK: %[[VAL_339:.*]] = udiv i32 %[[VAL_334]], 1024 |
| // CHECK: %[[VAL_340:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_341:.*]] = getelementptr inbounds float, float* %[[VAL_340]], i32 %[[VAL_334]] |
| // CHECK: %[[VAL_342:.*]] = load float, float* %[[VAL_341]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_342]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_343:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_343]], float* %[[VAL_26]], float* %[[VAL_343]]) |
| // CHECK: %[[VAL_344:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_345:.*]] = getelementptr inbounds float, float* %[[VAL_344]], i32 %[[VAL_334]] |
| // CHECK: %[[VAL_346:.*]] = load float, float* %[[VAL_345]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_346]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_347:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_347]], float* %[[VAL_24]], float* %[[VAL_347]]) |
| // CHECK: %[[VAL_348:.*]] = add i32 320, %[[VAL_69]] |
| // CHECK: %[[VAL_349:.*]] = add i32 %[[VAL_70]], 320 |
| // CHECK: %[[VAL_350:.*]] = mul nuw nsw i32 %[[VAL_349]], 1 |
| // CHECK: %[[VAL_351:.*]] = add nuw nsw i32 0, %[[VAL_350]] |
| // CHECK: %[[VAL_352:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_353:.*]] = add nuw nsw i32 %[[VAL_351]], %[[VAL_352]] |
| // CHECK: %[[VAL_354:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_355:.*]] = add nuw nsw i32 %[[VAL_353]], %[[VAL_354]] |
| // CHECK: %[[VAL_356:.*]] = udiv i32 %[[VAL_355]], 1 |
| // CHECK: %[[VAL_357:.*]] = urem i32 %[[VAL_356]], 32 |
| // CHECK: %[[VAL_358:.*]] = udiv i32 %[[VAL_355]], 32 |
| // CHECK: %[[VAL_359:.*]] = urem i32 %[[VAL_358]], 32 |
| // CHECK: %[[VAL_360:.*]] = udiv i32 %[[VAL_355]], 1024 |
| // CHECK: %[[VAL_361:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_362:.*]] = getelementptr inbounds float, float* %[[VAL_361]], i32 %[[VAL_355]] |
| // CHECK: %[[VAL_363:.*]] = load float, float* %[[VAL_362]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_363]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_364:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_364]], float* %[[VAL_26]], float* %[[VAL_364]]) |
| // CHECK: %[[VAL_365:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_366:.*]] = getelementptr inbounds float, float* %[[VAL_365]], i32 %[[VAL_355]] |
| // CHECK: %[[VAL_367:.*]] = load float, float* %[[VAL_366]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_367]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_368:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_368]], float* %[[VAL_24]], float* %[[VAL_368]]) |
| // CHECK: %[[VAL_369:.*]] = add i32 321, %[[VAL_69]] |
| // CHECK: %[[VAL_370:.*]] = add i32 %[[VAL_70]], 321 |
| // CHECK: %[[VAL_371:.*]] = mul nuw nsw i32 %[[VAL_370]], 1 |
| // CHECK: %[[VAL_372:.*]] = add nuw nsw i32 0, %[[VAL_371]] |
| // CHECK: %[[VAL_373:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_374:.*]] = add nuw nsw i32 %[[VAL_372]], %[[VAL_373]] |
| // CHECK: %[[VAL_375:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_376:.*]] = add nuw nsw i32 %[[VAL_374]], %[[VAL_375]] |
| // CHECK: %[[VAL_377:.*]] = udiv i32 %[[VAL_376]], 1 |
| // CHECK: %[[VAL_378:.*]] = urem i32 %[[VAL_377]], 32 |
| // CHECK: %[[VAL_379:.*]] = udiv i32 %[[VAL_376]], 32 |
| // CHECK: %[[VAL_380:.*]] = urem i32 %[[VAL_379]], 32 |
| // CHECK: %[[VAL_381:.*]] = udiv i32 %[[VAL_376]], 1024 |
| // CHECK: %[[VAL_382:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_383:.*]] = getelementptr inbounds float, float* %[[VAL_382]], i32 %[[VAL_376]] |
| // CHECK: %[[VAL_384:.*]] = load float, float* %[[VAL_383]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_384]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_385:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_385]], float* %[[VAL_26]], float* %[[VAL_385]]) |
| // CHECK: %[[VAL_386:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_387:.*]] = getelementptr inbounds float, float* %[[VAL_386]], i32 %[[VAL_376]] |
| // CHECK: %[[VAL_388:.*]] = load float, float* %[[VAL_387]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_388]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_389:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_389]], float* %[[VAL_24]], float* %[[VAL_389]]) |
| // CHECK: %[[VAL_390:.*]] = add i32 384, %[[VAL_69]] |
| // CHECK: %[[VAL_391:.*]] = add i32 %[[VAL_70]], 384 |
| // CHECK: %[[VAL_392:.*]] = mul nuw nsw i32 %[[VAL_391]], 1 |
| // CHECK: %[[VAL_393:.*]] = add nuw nsw i32 0, %[[VAL_392]] |
| // CHECK: %[[VAL_394:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_395:.*]] = add nuw nsw i32 %[[VAL_393]], %[[VAL_394]] |
| // CHECK: %[[VAL_396:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_397:.*]] = add nuw nsw i32 %[[VAL_395]], %[[VAL_396]] |
| // CHECK: %[[VAL_398:.*]] = udiv i32 %[[VAL_397]], 1 |
| // CHECK: %[[VAL_399:.*]] = urem i32 %[[VAL_398]], 32 |
| // CHECK: %[[VAL_400:.*]] = udiv i32 %[[VAL_397]], 32 |
| // CHECK: %[[VAL_401:.*]] = urem i32 %[[VAL_400]], 32 |
| // CHECK: %[[VAL_402:.*]] = udiv i32 %[[VAL_397]], 1024 |
| // CHECK: %[[VAL_403:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_404:.*]] = getelementptr inbounds float, float* %[[VAL_403]], i32 %[[VAL_397]] |
| // CHECK: %[[VAL_405:.*]] = load float, float* %[[VAL_404]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_405]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_406:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_406]], float* %[[VAL_26]], float* %[[VAL_406]]) |
| // CHECK: %[[VAL_407:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_408:.*]] = getelementptr inbounds float, float* %[[VAL_407]], i32 %[[VAL_397]] |
| // CHECK: %[[VAL_409:.*]] = load float, float* %[[VAL_408]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_409]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_410:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_410]], float* %[[VAL_24]], float* %[[VAL_410]]) |
| // CHECK: %[[VAL_411:.*]] = add i32 385, %[[VAL_69]] |
| // CHECK: %[[VAL_412:.*]] = add i32 %[[VAL_70]], 385 |
| // CHECK: %[[VAL_413:.*]] = mul nuw nsw i32 %[[VAL_412]], 1 |
| // CHECK: %[[VAL_414:.*]] = add nuw nsw i32 0, %[[VAL_413]] |
| // CHECK: %[[VAL_415:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_416:.*]] = add nuw nsw i32 %[[VAL_414]], %[[VAL_415]] |
| // CHECK: %[[VAL_417:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_418:.*]] = add nuw nsw i32 %[[VAL_416]], %[[VAL_417]] |
| // CHECK: %[[VAL_419:.*]] = udiv i32 %[[VAL_418]], 1 |
| // CHECK: %[[VAL_420:.*]] = urem i32 %[[VAL_419]], 32 |
| // CHECK: %[[VAL_421:.*]] = udiv i32 %[[VAL_418]], 32 |
| // CHECK: %[[VAL_422:.*]] = urem i32 %[[VAL_421]], 32 |
| // CHECK: %[[VAL_423:.*]] = udiv i32 %[[VAL_418]], 1024 |
| // CHECK: %[[VAL_424:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_425:.*]] = getelementptr inbounds float, float* %[[VAL_424]], i32 %[[VAL_418]] |
| // CHECK: %[[VAL_426:.*]] = load float, float* %[[VAL_425]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_426]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_427:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_427]], float* %[[VAL_26]], float* %[[VAL_427]]) |
| // CHECK: %[[VAL_428:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_429:.*]] = getelementptr inbounds float, float* %[[VAL_428]], i32 %[[VAL_418]] |
| // CHECK: %[[VAL_430:.*]] = load float, float* %[[VAL_429]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_430]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_431:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_431]], float* %[[VAL_24]], float* %[[VAL_431]]) |
| // CHECK: %[[VAL_432:.*]] = add i32 448, %[[VAL_69]] |
| // CHECK: %[[VAL_433:.*]] = add i32 %[[VAL_70]], 448 |
| // CHECK: %[[VAL_434:.*]] = mul nuw nsw i32 %[[VAL_433]], 1 |
| // CHECK: %[[VAL_435:.*]] = add nuw nsw i32 0, %[[VAL_434]] |
| // CHECK: %[[VAL_436:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_437:.*]] = add nuw nsw i32 %[[VAL_435]], %[[VAL_436]] |
| // CHECK: %[[VAL_438:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_439:.*]] = add nuw nsw i32 %[[VAL_437]], %[[VAL_438]] |
| // CHECK: %[[VAL_440:.*]] = udiv i32 %[[VAL_439]], 1 |
| // CHECK: %[[VAL_441:.*]] = urem i32 %[[VAL_440]], 32 |
| // CHECK: %[[VAL_442:.*]] = udiv i32 %[[VAL_439]], 32 |
| // CHECK: %[[VAL_443:.*]] = urem i32 %[[VAL_442]], 32 |
| // CHECK: %[[VAL_444:.*]] = udiv i32 %[[VAL_439]], 1024 |
| // CHECK: %[[VAL_445:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_446:.*]] = getelementptr inbounds float, float* %[[VAL_445]], i32 %[[VAL_439]] |
| // CHECK: %[[VAL_447:.*]] = load float, float* %[[VAL_446]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_447]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_448:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_448]], float* %[[VAL_26]], float* %[[VAL_448]]) |
| // CHECK: %[[VAL_449:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_450:.*]] = getelementptr inbounds float, float* %[[VAL_449]], i32 %[[VAL_439]] |
| // CHECK: %[[VAL_451:.*]] = load float, float* %[[VAL_450]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_451]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_452:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_452]], float* %[[VAL_24]], float* %[[VAL_452]]) |
| // CHECK: %[[VAL_453:.*]] = add i32 449, %[[VAL_69]] |
| // CHECK: %[[VAL_454:.*]] = add i32 %[[VAL_70]], 449 |
| // CHECK: %[[VAL_455:.*]] = mul nuw nsw i32 %[[VAL_454]], 1 |
| // CHECK: %[[VAL_456:.*]] = add nuw nsw i32 0, %[[VAL_455]] |
| // CHECK: %[[VAL_457:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_458:.*]] = add nuw nsw i32 %[[VAL_456]], %[[VAL_457]] |
| // CHECK: %[[VAL_459:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_460:.*]] = add nuw nsw i32 %[[VAL_458]], %[[VAL_459]] |
| // CHECK: %[[VAL_461:.*]] = udiv i32 %[[VAL_460]], 1 |
| // CHECK: %[[VAL_462:.*]] = urem i32 %[[VAL_461]], 32 |
| // CHECK: %[[VAL_463:.*]] = udiv i32 %[[VAL_460]], 32 |
| // CHECK: %[[VAL_464:.*]] = urem i32 %[[VAL_463]], 32 |
| // CHECK: %[[VAL_465:.*]] = udiv i32 %[[VAL_460]], 1024 |
| // CHECK: %[[VAL_466:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_467:.*]] = getelementptr inbounds float, float* %[[VAL_466]], i32 %[[VAL_460]] |
| // CHECK: %[[VAL_468:.*]] = load float, float* %[[VAL_467]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_468]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_469:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_469]], float* %[[VAL_26]], float* %[[VAL_469]]) |
| // CHECK: %[[VAL_470:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_471:.*]] = getelementptr inbounds float, float* %[[VAL_470]], i32 %[[VAL_460]] |
| // CHECK: %[[VAL_472:.*]] = load float, float* %[[VAL_471]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_472]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_473:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_473]], float* %[[VAL_24]], float* %[[VAL_473]]) |
| // CHECK: %[[VAL_474:.*]] = add i32 512, %[[VAL_69]] |
| // CHECK: %[[VAL_475:.*]] = add i32 %[[VAL_70]], 512 |
| // CHECK: %[[VAL_476:.*]] = mul nuw nsw i32 %[[VAL_475]], 1 |
| // CHECK: %[[VAL_477:.*]] = add nuw nsw i32 0, %[[VAL_476]] |
| // CHECK: %[[VAL_478:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_479:.*]] = add nuw nsw i32 %[[VAL_477]], %[[VAL_478]] |
| // CHECK: %[[VAL_480:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_481:.*]] = add nuw nsw i32 %[[VAL_479]], %[[VAL_480]] |
| // CHECK: %[[VAL_482:.*]] = udiv i32 %[[VAL_481]], 1 |
| // CHECK: %[[VAL_483:.*]] = urem i32 %[[VAL_482]], 32 |
| // CHECK: %[[VAL_484:.*]] = udiv i32 %[[VAL_481]], 32 |
| // CHECK: %[[VAL_485:.*]] = urem i32 %[[VAL_484]], 32 |
| // CHECK: %[[VAL_486:.*]] = udiv i32 %[[VAL_481]], 1024 |
| // CHECK: %[[VAL_487:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_488:.*]] = getelementptr inbounds float, float* %[[VAL_487]], i32 %[[VAL_481]] |
| // CHECK: %[[VAL_489:.*]] = load float, float* %[[VAL_488]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_489]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_490:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_490]], float* %[[VAL_26]], float* %[[VAL_490]]) |
| // CHECK: %[[VAL_491:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_492:.*]] = getelementptr inbounds float, float* %[[VAL_491]], i32 %[[VAL_481]] |
| // CHECK: %[[VAL_493:.*]] = load float, float* %[[VAL_492]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_493]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_494:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_494]], float* %[[VAL_24]], float* %[[VAL_494]]) |
| // CHECK: %[[VAL_495:.*]] = add i32 513, %[[VAL_69]] |
| // CHECK: %[[VAL_496:.*]] = add i32 %[[VAL_70]], 513 |
| // CHECK: %[[VAL_497:.*]] = mul nuw nsw i32 %[[VAL_496]], 1 |
| // CHECK: %[[VAL_498:.*]] = add nuw nsw i32 0, %[[VAL_497]] |
| // CHECK: %[[VAL_499:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_500:.*]] = add nuw nsw i32 %[[VAL_498]], %[[VAL_499]] |
| // CHECK: %[[VAL_501:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_502:.*]] = add nuw nsw i32 %[[VAL_500]], %[[VAL_501]] |
| // CHECK: %[[VAL_503:.*]] = udiv i32 %[[VAL_502]], 1 |
| // CHECK: %[[VAL_504:.*]] = urem i32 %[[VAL_503]], 32 |
| // CHECK: %[[VAL_505:.*]] = udiv i32 %[[VAL_502]], 32 |
| // CHECK: %[[VAL_506:.*]] = urem i32 %[[VAL_505]], 32 |
| // CHECK: %[[VAL_507:.*]] = udiv i32 %[[VAL_502]], 1024 |
| // CHECK: %[[VAL_508:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_509:.*]] = getelementptr inbounds float, float* %[[VAL_508]], i32 %[[VAL_502]] |
| // CHECK: %[[VAL_510:.*]] = load float, float* %[[VAL_509]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_510]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_511:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_511]], float* %[[VAL_26]], float* %[[VAL_511]]) |
| // CHECK: %[[VAL_512:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_513:.*]] = getelementptr inbounds float, float* %[[VAL_512]], i32 %[[VAL_502]] |
| // CHECK: %[[VAL_514:.*]] = load float, float* %[[VAL_513]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_514]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_515:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_515]], float* %[[VAL_24]], float* %[[VAL_515]]) |
| // CHECK: %[[VAL_516:.*]] = add i32 576, %[[VAL_69]] |
| // CHECK: %[[VAL_517:.*]] = add i32 %[[VAL_70]], 576 |
| // CHECK: %[[VAL_518:.*]] = mul nuw nsw i32 %[[VAL_517]], 1 |
| // CHECK: %[[VAL_519:.*]] = add nuw nsw i32 0, %[[VAL_518]] |
| // CHECK: %[[VAL_520:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_521:.*]] = add nuw nsw i32 %[[VAL_519]], %[[VAL_520]] |
| // CHECK: %[[VAL_522:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_523:.*]] = add nuw nsw i32 %[[VAL_521]], %[[VAL_522]] |
| // CHECK: %[[VAL_524:.*]] = udiv i32 %[[VAL_523]], 1 |
| // CHECK: %[[VAL_525:.*]] = urem i32 %[[VAL_524]], 32 |
| // CHECK: %[[VAL_526:.*]] = udiv i32 %[[VAL_523]], 32 |
| // CHECK: %[[VAL_527:.*]] = urem i32 %[[VAL_526]], 32 |
| // CHECK: %[[VAL_528:.*]] = udiv i32 %[[VAL_523]], 1024 |
| // CHECK: %[[VAL_529:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_530:.*]] = getelementptr inbounds float, float* %[[VAL_529]], i32 %[[VAL_523]] |
| // CHECK: %[[VAL_531:.*]] = load float, float* %[[VAL_530]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_531]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_532:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_532]], float* %[[VAL_26]], float* %[[VAL_532]]) |
| // CHECK: %[[VAL_533:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_534:.*]] = getelementptr inbounds float, float* %[[VAL_533]], i32 %[[VAL_523]] |
| // CHECK: %[[VAL_535:.*]] = load float, float* %[[VAL_534]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_535]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_536:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_536]], float* %[[VAL_24]], float* %[[VAL_536]]) |
| // CHECK: %[[VAL_537:.*]] = add i32 577, %[[VAL_69]] |
| // CHECK: %[[VAL_538:.*]] = add i32 %[[VAL_70]], 577 |
| // CHECK: %[[VAL_539:.*]] = mul nuw nsw i32 %[[VAL_538]], 1 |
| // CHECK: %[[VAL_540:.*]] = add nuw nsw i32 0, %[[VAL_539]] |
| // CHECK: %[[VAL_541:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_542:.*]] = add nuw nsw i32 %[[VAL_540]], %[[VAL_541]] |
| // CHECK: %[[VAL_543:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_544:.*]] = add nuw nsw i32 %[[VAL_542]], %[[VAL_543]] |
| // CHECK: %[[VAL_545:.*]] = udiv i32 %[[VAL_544]], 1 |
| // CHECK: %[[VAL_546:.*]] = urem i32 %[[VAL_545]], 32 |
| // CHECK: %[[VAL_547:.*]] = udiv i32 %[[VAL_544]], 32 |
| // CHECK: %[[VAL_548:.*]] = urem i32 %[[VAL_547]], 32 |
| // CHECK: %[[VAL_549:.*]] = udiv i32 %[[VAL_544]], 1024 |
| // CHECK: %[[VAL_550:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_551:.*]] = getelementptr inbounds float, float* %[[VAL_550]], i32 %[[VAL_544]] |
| // CHECK: %[[VAL_552:.*]] = load float, float* %[[VAL_551]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_552]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_553:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_553]], float* %[[VAL_26]], float* %[[VAL_553]]) |
| // CHECK: %[[VAL_554:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_555:.*]] = getelementptr inbounds float, float* %[[VAL_554]], i32 %[[VAL_544]] |
| // CHECK: %[[VAL_556:.*]] = load float, float* %[[VAL_555]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_556]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_557:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_557]], float* %[[VAL_24]], float* %[[VAL_557]]) |
| // CHECK: %[[VAL_558:.*]] = add i32 640, %[[VAL_69]] |
| // CHECK: %[[VAL_559:.*]] = add i32 %[[VAL_70]], 640 |
| // CHECK: %[[VAL_560:.*]] = mul nuw nsw i32 %[[VAL_559]], 1 |
| // CHECK: %[[VAL_561:.*]] = add nuw nsw i32 0, %[[VAL_560]] |
| // CHECK: %[[VAL_562:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_563:.*]] = add nuw nsw i32 %[[VAL_561]], %[[VAL_562]] |
| // CHECK: %[[VAL_564:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_565:.*]] = add nuw nsw i32 %[[VAL_563]], %[[VAL_564]] |
| // CHECK: %[[VAL_566:.*]] = udiv i32 %[[VAL_565]], 1 |
| // CHECK: %[[VAL_567:.*]] = urem i32 %[[VAL_566]], 32 |
| // CHECK: %[[VAL_568:.*]] = udiv i32 %[[VAL_565]], 32 |
| // CHECK: %[[VAL_569:.*]] = urem i32 %[[VAL_568]], 32 |
| // CHECK: %[[VAL_570:.*]] = udiv i32 %[[VAL_565]], 1024 |
| // CHECK: %[[VAL_571:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_572:.*]] = getelementptr inbounds float, float* %[[VAL_571]], i32 %[[VAL_565]] |
| // CHECK: %[[VAL_573:.*]] = load float, float* %[[VAL_572]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_573]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_574:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_574]], float* %[[VAL_26]], float* %[[VAL_574]]) |
| // CHECK: %[[VAL_575:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_576:.*]] = getelementptr inbounds float, float* %[[VAL_575]], i32 %[[VAL_565]] |
| // CHECK: %[[VAL_577:.*]] = load float, float* %[[VAL_576]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_577]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_578:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_578]], float* %[[VAL_24]], float* %[[VAL_578]]) |
| // CHECK: %[[VAL_579:.*]] = add i32 641, %[[VAL_69]] |
| // CHECK: %[[VAL_580:.*]] = add i32 %[[VAL_70]], 641 |
| // CHECK: %[[VAL_581:.*]] = mul nuw nsw i32 %[[VAL_580]], 1 |
| // CHECK: %[[VAL_582:.*]] = add nuw nsw i32 0, %[[VAL_581]] |
| // CHECK: %[[VAL_583:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_584:.*]] = add nuw nsw i32 %[[VAL_582]], %[[VAL_583]] |
| // CHECK: %[[VAL_585:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_586:.*]] = add nuw nsw i32 %[[VAL_584]], %[[VAL_585]] |
| // CHECK: %[[VAL_587:.*]] = udiv i32 %[[VAL_586]], 1 |
| // CHECK: %[[VAL_588:.*]] = urem i32 %[[VAL_587]], 32 |
| // CHECK: %[[VAL_589:.*]] = udiv i32 %[[VAL_586]], 32 |
| // CHECK: %[[VAL_590:.*]] = urem i32 %[[VAL_589]], 32 |
| // CHECK: %[[VAL_591:.*]] = udiv i32 %[[VAL_586]], 1024 |
| // CHECK: %[[VAL_592:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_593:.*]] = getelementptr inbounds float, float* %[[VAL_592]], i32 %[[VAL_586]] |
| // CHECK: %[[VAL_594:.*]] = load float, float* %[[VAL_593]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_594]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_595:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_595]], float* %[[VAL_26]], float* %[[VAL_595]]) |
| // CHECK: %[[VAL_596:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_597:.*]] = getelementptr inbounds float, float* %[[VAL_596]], i32 %[[VAL_586]] |
| // CHECK: %[[VAL_598:.*]] = load float, float* %[[VAL_597]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_598]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_599:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_599]], float* %[[VAL_24]], float* %[[VAL_599]]) |
| // CHECK: %[[VAL_600:.*]] = add i32 704, %[[VAL_69]] |
| // CHECK: %[[VAL_601:.*]] = add i32 %[[VAL_70]], 704 |
| // CHECK: %[[VAL_602:.*]] = mul nuw nsw i32 %[[VAL_601]], 1 |
| // CHECK: %[[VAL_603:.*]] = add nuw nsw i32 0, %[[VAL_602]] |
| // CHECK: %[[VAL_604:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_605:.*]] = add nuw nsw i32 %[[VAL_603]], %[[VAL_604]] |
| // CHECK: %[[VAL_606:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_607:.*]] = add nuw nsw i32 %[[VAL_605]], %[[VAL_606]] |
| // CHECK: %[[VAL_608:.*]] = udiv i32 %[[VAL_607]], 1 |
| // CHECK: %[[VAL_609:.*]] = urem i32 %[[VAL_608]], 32 |
| // CHECK: %[[VAL_610:.*]] = udiv i32 %[[VAL_607]], 32 |
| // CHECK: %[[VAL_611:.*]] = urem i32 %[[VAL_610]], 32 |
| // CHECK: %[[VAL_612:.*]] = udiv i32 %[[VAL_607]], 1024 |
| // CHECK: %[[VAL_613:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_614:.*]] = getelementptr inbounds float, float* %[[VAL_613]], i32 %[[VAL_607]] |
| // CHECK: %[[VAL_615:.*]] = load float, float* %[[VAL_614]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_615]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_616:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_616]], float* %[[VAL_26]], float* %[[VAL_616]]) |
| // CHECK: %[[VAL_617:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_618:.*]] = getelementptr inbounds float, float* %[[VAL_617]], i32 %[[VAL_607]] |
| // CHECK: %[[VAL_619:.*]] = load float, float* %[[VAL_618]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_619]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_620:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_620]], float* %[[VAL_24]], float* %[[VAL_620]]) |
| // CHECK: %[[VAL_621:.*]] = add i32 705, %[[VAL_69]] |
| // CHECK: %[[VAL_622:.*]] = add i32 %[[VAL_70]], 705 |
| // CHECK: %[[VAL_623:.*]] = mul nuw nsw i32 %[[VAL_622]], 1 |
| // CHECK: %[[VAL_624:.*]] = add nuw nsw i32 0, %[[VAL_623]] |
| // CHECK: %[[VAL_625:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_626:.*]] = add nuw nsw i32 %[[VAL_624]], %[[VAL_625]] |
| // CHECK: %[[VAL_627:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_628:.*]] = add nuw nsw i32 %[[VAL_626]], %[[VAL_627]] |
| // CHECK: %[[VAL_629:.*]] = udiv i32 %[[VAL_628]], 1 |
| // CHECK: %[[VAL_630:.*]] = urem i32 %[[VAL_629]], 32 |
| // CHECK: %[[VAL_631:.*]] = udiv i32 %[[VAL_628]], 32 |
| // CHECK: %[[VAL_632:.*]] = urem i32 %[[VAL_631]], 32 |
| // CHECK: %[[VAL_633:.*]] = udiv i32 %[[VAL_628]], 1024 |
| // CHECK: %[[VAL_634:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_635:.*]] = getelementptr inbounds float, float* %[[VAL_634]], i32 %[[VAL_628]] |
| // CHECK: %[[VAL_636:.*]] = load float, float* %[[VAL_635]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_636]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_637:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_637]], float* %[[VAL_26]], float* %[[VAL_637]]) |
| // CHECK: %[[VAL_638:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_639:.*]] = getelementptr inbounds float, float* %[[VAL_638]], i32 %[[VAL_628]] |
| // CHECK: %[[VAL_640:.*]] = load float, float* %[[VAL_639]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_640]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_641:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_641]], float* %[[VAL_24]], float* %[[VAL_641]]) |
| // CHECK: %[[VAL_642:.*]] = add i32 768, %[[VAL_69]] |
| // CHECK: %[[VAL_643:.*]] = add i32 %[[VAL_70]], 768 |
| // CHECK: %[[VAL_644:.*]] = mul nuw nsw i32 %[[VAL_643]], 1 |
| // CHECK: %[[VAL_645:.*]] = add nuw nsw i32 0, %[[VAL_644]] |
| // CHECK: %[[VAL_646:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_647:.*]] = add nuw nsw i32 %[[VAL_645]], %[[VAL_646]] |
| // CHECK: %[[VAL_648:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_649:.*]] = add nuw nsw i32 %[[VAL_647]], %[[VAL_648]] |
| // CHECK: %[[VAL_650:.*]] = udiv i32 %[[VAL_649]], 1 |
| // CHECK: %[[VAL_651:.*]] = urem i32 %[[VAL_650]], 32 |
| // CHECK: %[[VAL_652:.*]] = udiv i32 %[[VAL_649]], 32 |
| // CHECK: %[[VAL_653:.*]] = urem i32 %[[VAL_652]], 32 |
| // CHECK: %[[VAL_654:.*]] = udiv i32 %[[VAL_649]], 1024 |
| // CHECK: %[[VAL_655:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_656:.*]] = getelementptr inbounds float, float* %[[VAL_655]], i32 %[[VAL_649]] |
| // CHECK: %[[VAL_657:.*]] = load float, float* %[[VAL_656]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_657]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_658:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_658]], float* %[[VAL_26]], float* %[[VAL_658]]) |
| // CHECK: %[[VAL_659:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_660:.*]] = getelementptr inbounds float, float* %[[VAL_659]], i32 %[[VAL_649]] |
| // CHECK: %[[VAL_661:.*]] = load float, float* %[[VAL_660]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_661]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_662:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_662]], float* %[[VAL_24]], float* %[[VAL_662]]) |
| // CHECK: %[[VAL_663:.*]] = add i32 769, %[[VAL_69]] |
| // CHECK: %[[VAL_664:.*]] = add i32 %[[VAL_70]], 769 |
| // CHECK: %[[VAL_665:.*]] = mul nuw nsw i32 %[[VAL_664]], 1 |
| // CHECK: %[[VAL_666:.*]] = add nuw nsw i32 0, %[[VAL_665]] |
| // CHECK: %[[VAL_667:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_668:.*]] = add nuw nsw i32 %[[VAL_666]], %[[VAL_667]] |
| // CHECK: %[[VAL_669:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_670:.*]] = add nuw nsw i32 %[[VAL_668]], %[[VAL_669]] |
| // CHECK: %[[VAL_671:.*]] = udiv i32 %[[VAL_670]], 1 |
| // CHECK: %[[VAL_672:.*]] = urem i32 %[[VAL_671]], 32 |
| // CHECK: %[[VAL_673:.*]] = udiv i32 %[[VAL_670]], 32 |
| // CHECK: %[[VAL_674:.*]] = urem i32 %[[VAL_673]], 32 |
| // CHECK: %[[VAL_675:.*]] = udiv i32 %[[VAL_670]], 1024 |
| // CHECK: %[[VAL_676:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_677:.*]] = getelementptr inbounds float, float* %[[VAL_676]], i32 %[[VAL_670]] |
| // CHECK: %[[VAL_678:.*]] = load float, float* %[[VAL_677]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_678]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_679:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_679]], float* %[[VAL_26]], float* %[[VAL_679]]) |
| // CHECK: %[[VAL_680:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_681:.*]] = getelementptr inbounds float, float* %[[VAL_680]], i32 %[[VAL_670]] |
| // CHECK: %[[VAL_682:.*]] = load float, float* %[[VAL_681]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_682]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_683:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_683]], float* %[[VAL_24]], float* %[[VAL_683]]) |
| // CHECK: %[[VAL_684:.*]] = add i32 832, %[[VAL_69]] |
| // CHECK: %[[VAL_685:.*]] = add i32 %[[VAL_70]], 832 |
| // CHECK: %[[VAL_686:.*]] = mul nuw nsw i32 %[[VAL_685]], 1 |
| // CHECK: %[[VAL_687:.*]] = add nuw nsw i32 0, %[[VAL_686]] |
| // CHECK: %[[VAL_688:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_689:.*]] = add nuw nsw i32 %[[VAL_687]], %[[VAL_688]] |
| // CHECK: %[[VAL_690:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_691:.*]] = add nuw nsw i32 %[[VAL_689]], %[[VAL_690]] |
| // CHECK: %[[VAL_692:.*]] = udiv i32 %[[VAL_691]], 1 |
| // CHECK: %[[VAL_693:.*]] = urem i32 %[[VAL_692]], 32 |
| // CHECK: %[[VAL_694:.*]] = udiv i32 %[[VAL_691]], 32 |
| // CHECK: %[[VAL_695:.*]] = urem i32 %[[VAL_694]], 32 |
| // CHECK: %[[VAL_696:.*]] = udiv i32 %[[VAL_691]], 1024 |
| // CHECK: %[[VAL_697:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_698:.*]] = getelementptr inbounds float, float* %[[VAL_697]], i32 %[[VAL_691]] |
| // CHECK: %[[VAL_699:.*]] = load float, float* %[[VAL_698]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_699]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_700:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_700]], float* %[[VAL_26]], float* %[[VAL_700]]) |
| // CHECK: %[[VAL_701:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_702:.*]] = getelementptr inbounds float, float* %[[VAL_701]], i32 %[[VAL_691]] |
| // CHECK: %[[VAL_703:.*]] = load float, float* %[[VAL_702]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_703]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_704:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_704]], float* %[[VAL_24]], float* %[[VAL_704]]) |
| // CHECK: %[[VAL_705:.*]] = add i32 833, %[[VAL_69]] |
| // CHECK: %[[VAL_706:.*]] = add i32 %[[VAL_70]], 833 |
| // CHECK: %[[VAL_707:.*]] = mul nuw nsw i32 %[[VAL_706]], 1 |
| // CHECK: %[[VAL_708:.*]] = add nuw nsw i32 0, %[[VAL_707]] |
| // CHECK: %[[VAL_709:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_710:.*]] = add nuw nsw i32 %[[VAL_708]], %[[VAL_709]] |
| // CHECK: %[[VAL_711:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_712:.*]] = add nuw nsw i32 %[[VAL_710]], %[[VAL_711]] |
| // CHECK: %[[VAL_713:.*]] = udiv i32 %[[VAL_712]], 1 |
| // CHECK: %[[VAL_714:.*]] = urem i32 %[[VAL_713]], 32 |
| // CHECK: %[[VAL_715:.*]] = udiv i32 %[[VAL_712]], 32 |
| // CHECK: %[[VAL_716:.*]] = urem i32 %[[VAL_715]], 32 |
| // CHECK: %[[VAL_717:.*]] = udiv i32 %[[VAL_712]], 1024 |
| // CHECK: %[[VAL_718:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_719:.*]] = getelementptr inbounds float, float* %[[VAL_718]], i32 %[[VAL_712]] |
| // CHECK: %[[VAL_720:.*]] = load float, float* %[[VAL_719]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_720]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_721:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_721]], float* %[[VAL_26]], float* %[[VAL_721]]) |
| // CHECK: %[[VAL_722:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_723:.*]] = getelementptr inbounds float, float* %[[VAL_722]], i32 %[[VAL_712]] |
| // CHECK: %[[VAL_724:.*]] = load float, float* %[[VAL_723]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_724]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_725:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_725]], float* %[[VAL_24]], float* %[[VAL_725]]) |
| // CHECK: %[[VAL_726:.*]] = add i32 896, %[[VAL_69]] |
| // CHECK: %[[VAL_727:.*]] = add i32 %[[VAL_70]], 896 |
| // CHECK: %[[VAL_728:.*]] = mul nuw nsw i32 %[[VAL_727]], 1 |
| // CHECK: %[[VAL_729:.*]] = add nuw nsw i32 0, %[[VAL_728]] |
| // CHECK: %[[VAL_730:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_731:.*]] = add nuw nsw i32 %[[VAL_729]], %[[VAL_730]] |
| // CHECK: %[[VAL_732:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_733:.*]] = add nuw nsw i32 %[[VAL_731]], %[[VAL_732]] |
| // CHECK: %[[VAL_734:.*]] = udiv i32 %[[VAL_733]], 1 |
| // CHECK: %[[VAL_735:.*]] = urem i32 %[[VAL_734]], 32 |
| // CHECK: %[[VAL_736:.*]] = udiv i32 %[[VAL_733]], 32 |
| // CHECK: %[[VAL_737:.*]] = urem i32 %[[VAL_736]], 32 |
| // CHECK: %[[VAL_738:.*]] = udiv i32 %[[VAL_733]], 1024 |
| // CHECK: %[[VAL_739:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_740:.*]] = getelementptr inbounds float, float* %[[VAL_739]], i32 %[[VAL_733]] |
| // CHECK: %[[VAL_741:.*]] = load float, float* %[[VAL_740]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_741]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_742:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_742]], float* %[[VAL_26]], float* %[[VAL_742]]) |
| // CHECK: %[[VAL_743:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_744:.*]] = getelementptr inbounds float, float* %[[VAL_743]], i32 %[[VAL_733]] |
| // CHECK: %[[VAL_745:.*]] = load float, float* %[[VAL_744]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_745]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_746:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_746]], float* %[[VAL_24]], float* %[[VAL_746]]) |
| // CHECK: %[[VAL_747:.*]] = add i32 897, %[[VAL_69]] |
| // CHECK: %[[VAL_748:.*]] = add i32 %[[VAL_70]], 897 |
| // CHECK: %[[VAL_749:.*]] = mul nuw nsw i32 %[[VAL_748]], 1 |
| // CHECK: %[[VAL_750:.*]] = add nuw nsw i32 0, %[[VAL_749]] |
| // CHECK: %[[VAL_751:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_752:.*]] = add nuw nsw i32 %[[VAL_750]], %[[VAL_751]] |
| // CHECK: %[[VAL_753:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_754:.*]] = add nuw nsw i32 %[[VAL_752]], %[[VAL_753]] |
| // CHECK: %[[VAL_755:.*]] = udiv i32 %[[VAL_754]], 1 |
| // CHECK: %[[VAL_756:.*]] = urem i32 %[[VAL_755]], 32 |
| // CHECK: %[[VAL_757:.*]] = udiv i32 %[[VAL_754]], 32 |
| // CHECK: %[[VAL_758:.*]] = urem i32 %[[VAL_757]], 32 |
| // CHECK: %[[VAL_759:.*]] = udiv i32 %[[VAL_754]], 1024 |
| // CHECK: %[[VAL_760:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_761:.*]] = getelementptr inbounds float, float* %[[VAL_760]], i32 %[[VAL_754]] |
| // CHECK: %[[VAL_762:.*]] = load float, float* %[[VAL_761]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_762]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_763:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_763]], float* %[[VAL_26]], float* %[[VAL_763]]) |
| // CHECK: %[[VAL_764:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_765:.*]] = getelementptr inbounds float, float* %[[VAL_764]], i32 %[[VAL_754]] |
| // CHECK: %[[VAL_766:.*]] = load float, float* %[[VAL_765]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_766]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_767:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_767]], float* %[[VAL_24]], float* %[[VAL_767]]) |
| // CHECK: %[[VAL_768:.*]] = add i32 960, %[[VAL_69]] |
| // CHECK: %[[VAL_769:.*]] = add i32 %[[VAL_70]], 960 |
| // CHECK: %[[VAL_770:.*]] = mul nuw nsw i32 %[[VAL_769]], 1 |
| // CHECK: %[[VAL_771:.*]] = add nuw nsw i32 0, %[[VAL_770]] |
| // CHECK: %[[VAL_772:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_773:.*]] = add nuw nsw i32 %[[VAL_771]], %[[VAL_772]] |
| // CHECK: %[[VAL_774:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_775:.*]] = add nuw nsw i32 %[[VAL_773]], %[[VAL_774]] |
| // CHECK: %[[VAL_776:.*]] = udiv i32 %[[VAL_775]], 1 |
| // CHECK: %[[VAL_777:.*]] = urem i32 %[[VAL_776]], 32 |
| // CHECK: %[[VAL_778:.*]] = udiv i32 %[[VAL_775]], 32 |
| // CHECK: %[[VAL_779:.*]] = urem i32 %[[VAL_778]], 32 |
| // CHECK: %[[VAL_780:.*]] = udiv i32 %[[VAL_775]], 1024 |
| // CHECK: %[[VAL_781:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_782:.*]] = getelementptr inbounds float, float* %[[VAL_781]], i32 %[[VAL_775]] |
| // CHECK: %[[VAL_783:.*]] = load float, float* %[[VAL_782]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_783]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_784:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_784]], float* %[[VAL_26]], float* %[[VAL_784]]) |
| // CHECK: %[[VAL_785:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_786:.*]] = getelementptr inbounds float, float* %[[VAL_785]], i32 %[[VAL_775]] |
| // CHECK: %[[VAL_787:.*]] = load float, float* %[[VAL_786]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_787]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_788:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_788]], float* %[[VAL_24]], float* %[[VAL_788]]) |
| // CHECK: %[[VAL_789:.*]] = add i32 961, %[[VAL_69]] |
| // CHECK: %[[VAL_790:.*]] = add i32 %[[VAL_70]], 961 |
| // CHECK: %[[VAL_791:.*]] = mul nuw nsw i32 %[[VAL_790]], 1 |
| // CHECK: %[[VAL_792:.*]] = add nuw nsw i32 0, %[[VAL_791]] |
| // CHECK: %[[VAL_793:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_794:.*]] = add nuw nsw i32 %[[VAL_792]], %[[VAL_793]] |
| // CHECK: %[[VAL_795:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_796:.*]] = add nuw nsw i32 %[[VAL_794]], %[[VAL_795]] |
| // CHECK: %[[VAL_797:.*]] = udiv i32 %[[VAL_796]], 1 |
| // CHECK: %[[VAL_798:.*]] = urem i32 %[[VAL_797]], 32 |
| // CHECK: %[[VAL_799:.*]] = udiv i32 %[[VAL_796]], 32 |
| // CHECK: %[[VAL_800:.*]] = urem i32 %[[VAL_799]], 32 |
| // CHECK: %[[VAL_801:.*]] = udiv i32 %[[VAL_796]], 1024 |
| // CHECK: %[[VAL_802:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_803:.*]] = getelementptr inbounds float, float* %[[VAL_802]], i32 %[[VAL_796]] |
| // CHECK: %[[VAL_804:.*]] = load float, float* %[[VAL_803]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_804]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_805:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_805]], float* %[[VAL_26]], float* %[[VAL_805]]) |
| // CHECK: %[[VAL_806:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_807:.*]] = getelementptr inbounds float, float* %[[VAL_806]], i32 %[[VAL_796]] |
| // CHECK: %[[VAL_808:.*]] = load float, float* %[[VAL_807]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_808]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_809:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_809]], float* %[[VAL_24]], float* %[[VAL_809]]) |
| // CHECK: %[[VAL_810:.*]] = add i32 1024, %[[VAL_69]] |
| // CHECK: %[[VAL_811:.*]] = add i32 %[[VAL_70]], 1024 |
| // CHECK: %[[VAL_812:.*]] = mul nuw nsw i32 %[[VAL_811]], 1 |
| // CHECK: %[[VAL_813:.*]] = add nuw nsw i32 0, %[[VAL_812]] |
| // CHECK: %[[VAL_814:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_815:.*]] = add nuw nsw i32 %[[VAL_813]], %[[VAL_814]] |
| // CHECK: %[[VAL_816:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_817:.*]] = add nuw nsw i32 %[[VAL_815]], %[[VAL_816]] |
| // CHECK: %[[VAL_818:.*]] = udiv i32 %[[VAL_817]], 1 |
| // CHECK: %[[VAL_819:.*]] = urem i32 %[[VAL_818]], 32 |
| // CHECK: %[[VAL_820:.*]] = udiv i32 %[[VAL_817]], 32 |
| // CHECK: %[[VAL_821:.*]] = urem i32 %[[VAL_820]], 32 |
| // CHECK: %[[VAL_822:.*]] = udiv i32 %[[VAL_817]], 1024 |
| // CHECK: %[[VAL_823:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_824:.*]] = getelementptr inbounds float, float* %[[VAL_823]], i32 %[[VAL_817]] |
| // CHECK: %[[VAL_825:.*]] = load float, float* %[[VAL_824]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_825]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_826:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_826]], float* %[[VAL_26]], float* %[[VAL_826]]) |
| // CHECK: %[[VAL_827:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_828:.*]] = getelementptr inbounds float, float* %[[VAL_827]], i32 %[[VAL_817]] |
| // CHECK: %[[VAL_829:.*]] = load float, float* %[[VAL_828]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_829]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_830:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_830]], float* %[[VAL_24]], float* %[[VAL_830]]) |
| // CHECK: %[[VAL_831:.*]] = add i32 1025, %[[VAL_69]] |
| // CHECK: %[[VAL_832:.*]] = add i32 %[[VAL_70]], 1025 |
| // CHECK: %[[VAL_833:.*]] = mul nuw nsw i32 %[[VAL_832]], 1 |
| // CHECK: %[[VAL_834:.*]] = add nuw nsw i32 0, %[[VAL_833]] |
| // CHECK: %[[VAL_835:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_836:.*]] = add nuw nsw i32 %[[VAL_834]], %[[VAL_835]] |
| // CHECK: %[[VAL_837:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_838:.*]] = add nuw nsw i32 %[[VAL_836]], %[[VAL_837]] |
| // CHECK: %[[VAL_839:.*]] = udiv i32 %[[VAL_838]], 1 |
| // CHECK: %[[VAL_840:.*]] = urem i32 %[[VAL_839]], 32 |
| // CHECK: %[[VAL_841:.*]] = udiv i32 %[[VAL_838]], 32 |
| // CHECK: %[[VAL_842:.*]] = urem i32 %[[VAL_841]], 32 |
| // CHECK: %[[VAL_843:.*]] = udiv i32 %[[VAL_838]], 1024 |
| // CHECK: %[[VAL_844:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_845:.*]] = getelementptr inbounds float, float* %[[VAL_844]], i32 %[[VAL_838]] |
| // CHECK: %[[VAL_846:.*]] = load float, float* %[[VAL_845]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_846]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_847:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_847]], float* %[[VAL_26]], float* %[[VAL_847]]) |
| // CHECK: %[[VAL_848:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_849:.*]] = getelementptr inbounds float, float* %[[VAL_848]], i32 %[[VAL_838]] |
| // CHECK: %[[VAL_850:.*]] = load float, float* %[[VAL_849]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_850]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_851:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_851]], float* %[[VAL_24]], float* %[[VAL_851]]) |
| // CHECK: %[[VAL_852:.*]] = add i32 1088, %[[VAL_69]] |
| // CHECK: %[[VAL_853:.*]] = add i32 %[[VAL_70]], 1088 |
| // CHECK: %[[VAL_854:.*]] = mul nuw nsw i32 %[[VAL_853]], 1 |
| // CHECK: %[[VAL_855:.*]] = add nuw nsw i32 0, %[[VAL_854]] |
| // CHECK: %[[VAL_856:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_857:.*]] = add nuw nsw i32 %[[VAL_855]], %[[VAL_856]] |
| // CHECK: %[[VAL_858:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_859:.*]] = add nuw nsw i32 %[[VAL_857]], %[[VAL_858]] |
| // CHECK: %[[VAL_860:.*]] = udiv i32 %[[VAL_859]], 1 |
| // CHECK: %[[VAL_861:.*]] = urem i32 %[[VAL_860]], 32 |
| // CHECK: %[[VAL_862:.*]] = udiv i32 %[[VAL_859]], 32 |
| // CHECK: %[[VAL_863:.*]] = urem i32 %[[VAL_862]], 32 |
| // CHECK: %[[VAL_864:.*]] = udiv i32 %[[VAL_859]], 1024 |
| // CHECK: %[[VAL_865:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_866:.*]] = getelementptr inbounds float, float* %[[VAL_865]], i32 %[[VAL_859]] |
| // CHECK: %[[VAL_867:.*]] = load float, float* %[[VAL_866]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_867]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_868:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_868]], float* %[[VAL_26]], float* %[[VAL_868]]) |
| // CHECK: %[[VAL_869:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_870:.*]] = getelementptr inbounds float, float* %[[VAL_869]], i32 %[[VAL_859]] |
| // CHECK: %[[VAL_871:.*]] = load float, float* %[[VAL_870]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_871]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_872:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_872]], float* %[[VAL_24]], float* %[[VAL_872]]) |
| // CHECK: %[[VAL_873:.*]] = add i32 1089, %[[VAL_69]] |
| // CHECK: %[[VAL_874:.*]] = add i32 %[[VAL_70]], 1089 |
| // CHECK: %[[VAL_875:.*]] = mul nuw nsw i32 %[[VAL_874]], 1 |
| // CHECK: %[[VAL_876:.*]] = add nuw nsw i32 0, %[[VAL_875]] |
| // CHECK: %[[VAL_877:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_878:.*]] = add nuw nsw i32 %[[VAL_876]], %[[VAL_877]] |
| // CHECK: %[[VAL_879:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_880:.*]] = add nuw nsw i32 %[[VAL_878]], %[[VAL_879]] |
| // CHECK: %[[VAL_881:.*]] = udiv i32 %[[VAL_880]], 1 |
| // CHECK: %[[VAL_882:.*]] = urem i32 %[[VAL_881]], 32 |
| // CHECK: %[[VAL_883:.*]] = udiv i32 %[[VAL_880]], 32 |
| // CHECK: %[[VAL_884:.*]] = urem i32 %[[VAL_883]], 32 |
| // CHECK: %[[VAL_885:.*]] = udiv i32 %[[VAL_880]], 1024 |
| // CHECK: %[[VAL_886:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_887:.*]] = getelementptr inbounds float, float* %[[VAL_886]], i32 %[[VAL_880]] |
| // CHECK: %[[VAL_888:.*]] = load float, float* %[[VAL_887]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_888]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_889:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_889]], float* %[[VAL_26]], float* %[[VAL_889]]) |
| // CHECK: %[[VAL_890:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_891:.*]] = getelementptr inbounds float, float* %[[VAL_890]], i32 %[[VAL_880]] |
| // CHECK: %[[VAL_892:.*]] = load float, float* %[[VAL_891]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_892]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_893:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_893]], float* %[[VAL_24]], float* %[[VAL_893]]) |
| // CHECK: %[[VAL_894:.*]] = add i32 1152, %[[VAL_69]] |
| // CHECK: %[[VAL_895:.*]] = add i32 %[[VAL_70]], 1152 |
| // CHECK: %[[VAL_896:.*]] = mul nuw nsw i32 %[[VAL_895]], 1 |
| // CHECK: %[[VAL_897:.*]] = add nuw nsw i32 0, %[[VAL_896]] |
| // CHECK: %[[VAL_898:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_899:.*]] = add nuw nsw i32 %[[VAL_897]], %[[VAL_898]] |
| // CHECK: %[[VAL_900:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_901:.*]] = add nuw nsw i32 %[[VAL_899]], %[[VAL_900]] |
| // CHECK: %[[VAL_902:.*]] = udiv i32 %[[VAL_901]], 1 |
| // CHECK: %[[VAL_903:.*]] = urem i32 %[[VAL_902]], 32 |
| // CHECK: %[[VAL_904:.*]] = udiv i32 %[[VAL_901]], 32 |
| // CHECK: %[[VAL_905:.*]] = urem i32 %[[VAL_904]], 32 |
| // CHECK: %[[VAL_906:.*]] = udiv i32 %[[VAL_901]], 1024 |
| // CHECK: %[[VAL_907:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_908:.*]] = getelementptr inbounds float, float* %[[VAL_907]], i32 %[[VAL_901]] |
| // CHECK: %[[VAL_909:.*]] = load float, float* %[[VAL_908]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_909]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_910:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_910]], float* %[[VAL_26]], float* %[[VAL_910]]) |
| // CHECK: %[[VAL_911:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_912:.*]] = getelementptr inbounds float, float* %[[VAL_911]], i32 %[[VAL_901]] |
| // CHECK: %[[VAL_913:.*]] = load float, float* %[[VAL_912]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_913]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_914:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_914]], float* %[[VAL_24]], float* %[[VAL_914]]) |
| // CHECK: %[[VAL_915:.*]] = add i32 1153, %[[VAL_69]] |
| // CHECK: %[[VAL_916:.*]] = add i32 %[[VAL_70]], 1153 |
| // CHECK: %[[VAL_917:.*]] = mul nuw nsw i32 %[[VAL_916]], 1 |
| // CHECK: %[[VAL_918:.*]] = add nuw nsw i32 0, %[[VAL_917]] |
| // CHECK: %[[VAL_919:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_920:.*]] = add nuw nsw i32 %[[VAL_918]], %[[VAL_919]] |
| // CHECK: %[[VAL_921:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_922:.*]] = add nuw nsw i32 %[[VAL_920]], %[[VAL_921]] |
| // CHECK: %[[VAL_923:.*]] = udiv i32 %[[VAL_922]], 1 |
| // CHECK: %[[VAL_924:.*]] = urem i32 %[[VAL_923]], 32 |
| // CHECK: %[[VAL_925:.*]] = udiv i32 %[[VAL_922]], 32 |
| // CHECK: %[[VAL_926:.*]] = urem i32 %[[VAL_925]], 32 |
| // CHECK: %[[VAL_927:.*]] = udiv i32 %[[VAL_922]], 1024 |
| // CHECK: %[[VAL_928:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_929:.*]] = getelementptr inbounds float, float* %[[VAL_928]], i32 %[[VAL_922]] |
| // CHECK: %[[VAL_930:.*]] = load float, float* %[[VAL_929]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_930]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_931:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_931]], float* %[[VAL_26]], float* %[[VAL_931]]) |
| // CHECK: %[[VAL_932:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_933:.*]] = getelementptr inbounds float, float* %[[VAL_932]], i32 %[[VAL_922]] |
| // CHECK: %[[VAL_934:.*]] = load float, float* %[[VAL_933]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_934]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_935:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_935]], float* %[[VAL_24]], float* %[[VAL_935]]) |
| // CHECK: %[[VAL_936:.*]] = add i32 1216, %[[VAL_69]] |
| // CHECK: %[[VAL_937:.*]] = add i32 %[[VAL_70]], 1216 |
| // CHECK: %[[VAL_938:.*]] = mul nuw nsw i32 %[[VAL_937]], 1 |
| // CHECK: %[[VAL_939:.*]] = add nuw nsw i32 0, %[[VAL_938]] |
| // CHECK: %[[VAL_940:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_941:.*]] = add nuw nsw i32 %[[VAL_939]], %[[VAL_940]] |
| // CHECK: %[[VAL_942:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_943:.*]] = add nuw nsw i32 %[[VAL_941]], %[[VAL_942]] |
| // CHECK: %[[VAL_944:.*]] = udiv i32 %[[VAL_943]], 1 |
| // CHECK: %[[VAL_945:.*]] = urem i32 %[[VAL_944]], 32 |
| // CHECK: %[[VAL_946:.*]] = udiv i32 %[[VAL_943]], 32 |
| // CHECK: %[[VAL_947:.*]] = urem i32 %[[VAL_946]], 32 |
| // CHECK: %[[VAL_948:.*]] = udiv i32 %[[VAL_943]], 1024 |
| // CHECK: %[[VAL_949:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_950:.*]] = getelementptr inbounds float, float* %[[VAL_949]], i32 %[[VAL_943]] |
| // CHECK: %[[VAL_951:.*]] = load float, float* %[[VAL_950]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_951]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_952:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_952]], float* %[[VAL_26]], float* %[[VAL_952]]) |
| // CHECK: %[[VAL_953:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_954:.*]] = getelementptr inbounds float, float* %[[VAL_953]], i32 %[[VAL_943]] |
| // CHECK: %[[VAL_955:.*]] = load float, float* %[[VAL_954]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_955]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_956:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_956]], float* %[[VAL_24]], float* %[[VAL_956]]) |
| // CHECK: %[[VAL_957:.*]] = add i32 1217, %[[VAL_69]] |
| // CHECK: %[[VAL_958:.*]] = add i32 %[[VAL_70]], 1217 |
| // CHECK: %[[VAL_959:.*]] = mul nuw nsw i32 %[[VAL_958]], 1 |
| // CHECK: %[[VAL_960:.*]] = add nuw nsw i32 0, %[[VAL_959]] |
| // CHECK: %[[VAL_961:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_962:.*]] = add nuw nsw i32 %[[VAL_960]], %[[VAL_961]] |
| // CHECK: %[[VAL_963:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_964:.*]] = add nuw nsw i32 %[[VAL_962]], %[[VAL_963]] |
| // CHECK: %[[VAL_965:.*]] = udiv i32 %[[VAL_964]], 1 |
| // CHECK: %[[VAL_966:.*]] = urem i32 %[[VAL_965]], 32 |
| // CHECK: %[[VAL_967:.*]] = udiv i32 %[[VAL_964]], 32 |
| // CHECK: %[[VAL_968:.*]] = urem i32 %[[VAL_967]], 32 |
| // CHECK: %[[VAL_969:.*]] = udiv i32 %[[VAL_964]], 1024 |
| // CHECK: %[[VAL_970:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_971:.*]] = getelementptr inbounds float, float* %[[VAL_970]], i32 %[[VAL_964]] |
| // CHECK: %[[VAL_972:.*]] = load float, float* %[[VAL_971]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_972]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_973:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_973]], float* %[[VAL_26]], float* %[[VAL_973]]) |
| // CHECK: %[[VAL_974:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_975:.*]] = getelementptr inbounds float, float* %[[VAL_974]], i32 %[[VAL_964]] |
| // CHECK: %[[VAL_976:.*]] = load float, float* %[[VAL_975]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_976]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_977:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_977]], float* %[[VAL_24]], float* %[[VAL_977]]) |
| // CHECK: %[[VAL_978:.*]] = add i32 1280, %[[VAL_69]] |
| // CHECK: %[[VAL_979:.*]] = add i32 %[[VAL_70]], 1280 |
| // CHECK: %[[VAL_980:.*]] = mul nuw nsw i32 %[[VAL_979]], 1 |
| // CHECK: %[[VAL_981:.*]] = add nuw nsw i32 0, %[[VAL_980]] |
| // CHECK: %[[VAL_982:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_983:.*]] = add nuw nsw i32 %[[VAL_981]], %[[VAL_982]] |
| // CHECK: %[[VAL_984:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_985:.*]] = add nuw nsw i32 %[[VAL_983]], %[[VAL_984]] |
| // CHECK: %[[VAL_986:.*]] = udiv i32 %[[VAL_985]], 1 |
| // CHECK: %[[VAL_987:.*]] = urem i32 %[[VAL_986]], 32 |
| // CHECK: %[[VAL_988:.*]] = udiv i32 %[[VAL_985]], 32 |
| // CHECK: %[[VAL_989:.*]] = urem i32 %[[VAL_988]], 32 |
| // CHECK: %[[VAL_990:.*]] = udiv i32 %[[VAL_985]], 1024 |
| // CHECK: %[[VAL_991:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_992:.*]] = getelementptr inbounds float, float* %[[VAL_991]], i32 %[[VAL_985]] |
| // CHECK: %[[VAL_993:.*]] = load float, float* %[[VAL_992]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_993]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_994:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_994]], float* %[[VAL_26]], float* %[[VAL_994]]) |
| // CHECK: %[[VAL_995:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_996:.*]] = getelementptr inbounds float, float* %[[VAL_995]], i32 %[[VAL_985]] |
| // CHECK: %[[VAL_997:.*]] = load float, float* %[[VAL_996]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_997]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_998:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_998]], float* %[[VAL_24]], float* %[[VAL_998]]) |
| // CHECK: %[[VAL_999:.*]] = add i32 1281, %[[VAL_69]] |
| // CHECK: %[[VAL_1000:.*]] = add i32 %[[VAL_70]], 1281 |
| // CHECK: %[[VAL_1001:.*]] = mul nuw nsw i32 %[[VAL_1000]], 1 |
| // CHECK: %[[VAL_1002:.*]] = add nuw nsw i32 0, %[[VAL_1001]] |
| // CHECK: %[[VAL_1003:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1004:.*]] = add nuw nsw i32 %[[VAL_1002]], %[[VAL_1003]] |
| // CHECK: %[[VAL_1005:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1006:.*]] = add nuw nsw i32 %[[VAL_1004]], %[[VAL_1005]] |
| // CHECK: %[[VAL_1007:.*]] = udiv i32 %[[VAL_1006]], 1 |
| // CHECK: %[[VAL_1008:.*]] = urem i32 %[[VAL_1007]], 32 |
| // CHECK: %[[VAL_1009:.*]] = udiv i32 %[[VAL_1006]], 32 |
| // CHECK: %[[VAL_1010:.*]] = urem i32 %[[VAL_1009]], 32 |
| // CHECK: %[[VAL_1011:.*]] = udiv i32 %[[VAL_1006]], 1024 |
| // CHECK: %[[VAL_1012:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1013:.*]] = getelementptr inbounds float, float* %[[VAL_1012]], i32 %[[VAL_1006]] |
| // CHECK: %[[VAL_1014:.*]] = load float, float* %[[VAL_1013]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1014]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1015:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1015]], float* %[[VAL_26]], float* %[[VAL_1015]]) |
| // CHECK: %[[VAL_1016:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1017:.*]] = getelementptr inbounds float, float* %[[VAL_1016]], i32 %[[VAL_1006]] |
| // CHECK: %[[VAL_1018:.*]] = load float, float* %[[VAL_1017]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1018]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1019:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1019]], float* %[[VAL_24]], float* %[[VAL_1019]]) |
| // CHECK: %[[VAL_1020:.*]] = add i32 1344, %[[VAL_69]] |
| // CHECK: %[[VAL_1021:.*]] = add i32 %[[VAL_70]], 1344 |
| // CHECK: %[[VAL_1022:.*]] = mul nuw nsw i32 %[[VAL_1021]], 1 |
| // CHECK: %[[VAL_1023:.*]] = add nuw nsw i32 0, %[[VAL_1022]] |
| // CHECK: %[[VAL_1024:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1025:.*]] = add nuw nsw i32 %[[VAL_1023]], %[[VAL_1024]] |
| // CHECK: %[[VAL_1026:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1027:.*]] = add nuw nsw i32 %[[VAL_1025]], %[[VAL_1026]] |
| // CHECK: %[[VAL_1028:.*]] = udiv i32 %[[VAL_1027]], 1 |
| // CHECK: %[[VAL_1029:.*]] = urem i32 %[[VAL_1028]], 32 |
| // CHECK: %[[VAL_1030:.*]] = udiv i32 %[[VAL_1027]], 32 |
| // CHECK: %[[VAL_1031:.*]] = urem i32 %[[VAL_1030]], 32 |
| // CHECK: %[[VAL_1032:.*]] = udiv i32 %[[VAL_1027]], 1024 |
| // CHECK: %[[VAL_1033:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1034:.*]] = getelementptr inbounds float, float* %[[VAL_1033]], i32 %[[VAL_1027]] |
| // CHECK: %[[VAL_1035:.*]] = load float, float* %[[VAL_1034]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1035]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1036:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1036]], float* %[[VAL_26]], float* %[[VAL_1036]]) |
| // CHECK: %[[VAL_1037:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1038:.*]] = getelementptr inbounds float, float* %[[VAL_1037]], i32 %[[VAL_1027]] |
| // CHECK: %[[VAL_1039:.*]] = load float, float* %[[VAL_1038]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1039]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1040:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1040]], float* %[[VAL_24]], float* %[[VAL_1040]]) |
| // CHECK: %[[VAL_1041:.*]] = add i32 1345, %[[VAL_69]] |
| // CHECK: %[[VAL_1042:.*]] = add i32 %[[VAL_70]], 1345 |
| // CHECK: %[[VAL_1043:.*]] = mul nuw nsw i32 %[[VAL_1042]], 1 |
| // CHECK: %[[VAL_1044:.*]] = add nuw nsw i32 0, %[[VAL_1043]] |
| // CHECK: %[[VAL_1045:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1046:.*]] = add nuw nsw i32 %[[VAL_1044]], %[[VAL_1045]] |
| // CHECK: %[[VAL_1047:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1048:.*]] = add nuw nsw i32 %[[VAL_1046]], %[[VAL_1047]] |
| // CHECK: %[[VAL_1049:.*]] = udiv i32 %[[VAL_1048]], 1 |
| // CHECK: %[[VAL_1050:.*]] = urem i32 %[[VAL_1049]], 32 |
| // CHECK: %[[VAL_1051:.*]] = udiv i32 %[[VAL_1048]], 32 |
| // CHECK: %[[VAL_1052:.*]] = urem i32 %[[VAL_1051]], 32 |
| // CHECK: %[[VAL_1053:.*]] = udiv i32 %[[VAL_1048]], 1024 |
| // CHECK: %[[VAL_1054:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1055:.*]] = getelementptr inbounds float, float* %[[VAL_1054]], i32 %[[VAL_1048]] |
| // CHECK: %[[VAL_1056:.*]] = load float, float* %[[VAL_1055]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1056]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1057:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1057]], float* %[[VAL_26]], float* %[[VAL_1057]]) |
| // CHECK: %[[VAL_1058:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1059:.*]] = getelementptr inbounds float, float* %[[VAL_1058]], i32 %[[VAL_1048]] |
| // CHECK: %[[VAL_1060:.*]] = load float, float* %[[VAL_1059]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1060]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1061:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1061]], float* %[[VAL_24]], float* %[[VAL_1061]]) |
| // CHECK: %[[VAL_1062:.*]] = add i32 1408, %[[VAL_69]] |
| // CHECK: %[[VAL_1063:.*]] = add i32 %[[VAL_70]], 1408 |
| // CHECK: %[[VAL_1064:.*]] = mul nuw nsw i32 %[[VAL_1063]], 1 |
| // CHECK: %[[VAL_1065:.*]] = add nuw nsw i32 0, %[[VAL_1064]] |
| // CHECK: %[[VAL_1066:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1067:.*]] = add nuw nsw i32 %[[VAL_1065]], %[[VAL_1066]] |
| // CHECK: %[[VAL_1068:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1069:.*]] = add nuw nsw i32 %[[VAL_1067]], %[[VAL_1068]] |
| // CHECK: %[[VAL_1070:.*]] = udiv i32 %[[VAL_1069]], 1 |
| // CHECK: %[[VAL_1071:.*]] = urem i32 %[[VAL_1070]], 32 |
| // CHECK: %[[VAL_1072:.*]] = udiv i32 %[[VAL_1069]], 32 |
| // CHECK: %[[VAL_1073:.*]] = urem i32 %[[VAL_1072]], 32 |
| // CHECK: %[[VAL_1074:.*]] = udiv i32 %[[VAL_1069]], 1024 |
| // CHECK: %[[VAL_1075:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1076:.*]] = getelementptr inbounds float, float* %[[VAL_1075]], i32 %[[VAL_1069]] |
| // CHECK: %[[VAL_1077:.*]] = load float, float* %[[VAL_1076]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1077]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1078:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1078]], float* %[[VAL_26]], float* %[[VAL_1078]]) |
| // CHECK: %[[VAL_1079:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1080:.*]] = getelementptr inbounds float, float* %[[VAL_1079]], i32 %[[VAL_1069]] |
| // CHECK: %[[VAL_1081:.*]] = load float, float* %[[VAL_1080]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1081]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1082:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1082]], float* %[[VAL_24]], float* %[[VAL_1082]]) |
| // CHECK: %[[VAL_1083:.*]] = add i32 1409, %[[VAL_69]] |
| // CHECK: %[[VAL_1084:.*]] = add i32 %[[VAL_70]], 1409 |
| // CHECK: %[[VAL_1085:.*]] = mul nuw nsw i32 %[[VAL_1084]], 1 |
| // CHECK: %[[VAL_1086:.*]] = add nuw nsw i32 0, %[[VAL_1085]] |
| // CHECK: %[[VAL_1087:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1088:.*]] = add nuw nsw i32 %[[VAL_1086]], %[[VAL_1087]] |
| // CHECK: %[[VAL_1089:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1090:.*]] = add nuw nsw i32 %[[VAL_1088]], %[[VAL_1089]] |
| // CHECK: %[[VAL_1091:.*]] = udiv i32 %[[VAL_1090]], 1 |
| // CHECK: %[[VAL_1092:.*]] = urem i32 %[[VAL_1091]], 32 |
| // CHECK: %[[VAL_1093:.*]] = udiv i32 %[[VAL_1090]], 32 |
| // CHECK: %[[VAL_1094:.*]] = urem i32 %[[VAL_1093]], 32 |
| // CHECK: %[[VAL_1095:.*]] = udiv i32 %[[VAL_1090]], 1024 |
| // CHECK: %[[VAL_1096:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1097:.*]] = getelementptr inbounds float, float* %[[VAL_1096]], i32 %[[VAL_1090]] |
| // CHECK: %[[VAL_1098:.*]] = load float, float* %[[VAL_1097]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1098]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1099:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1099]], float* %[[VAL_26]], float* %[[VAL_1099]]) |
| // CHECK: %[[VAL_1100:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1101:.*]] = getelementptr inbounds float, float* %[[VAL_1100]], i32 %[[VAL_1090]] |
| // CHECK: %[[VAL_1102:.*]] = load float, float* %[[VAL_1101]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1102]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1103:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1103]], float* %[[VAL_24]], float* %[[VAL_1103]]) |
| // CHECK: %[[VAL_1104:.*]] = add i32 1472, %[[VAL_69]] |
| // CHECK: %[[VAL_1105:.*]] = add i32 %[[VAL_70]], 1472 |
| // CHECK: %[[VAL_1106:.*]] = mul nuw nsw i32 %[[VAL_1105]], 1 |
| // CHECK: %[[VAL_1107:.*]] = add nuw nsw i32 0, %[[VAL_1106]] |
| // CHECK: %[[VAL_1108:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1109:.*]] = add nuw nsw i32 %[[VAL_1107]], %[[VAL_1108]] |
| // CHECK: %[[VAL_1110:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1111:.*]] = add nuw nsw i32 %[[VAL_1109]], %[[VAL_1110]] |
| // CHECK: %[[VAL_1112:.*]] = udiv i32 %[[VAL_1111]], 1 |
| // CHECK: %[[VAL_1113:.*]] = urem i32 %[[VAL_1112]], 32 |
| // CHECK: %[[VAL_1114:.*]] = udiv i32 %[[VAL_1111]], 32 |
| // CHECK: %[[VAL_1115:.*]] = urem i32 %[[VAL_1114]], 32 |
| // CHECK: %[[VAL_1116:.*]] = udiv i32 %[[VAL_1111]], 1024 |
| // CHECK: %[[VAL_1117:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1118:.*]] = getelementptr inbounds float, float* %[[VAL_1117]], i32 %[[VAL_1111]] |
| // CHECK: %[[VAL_1119:.*]] = load float, float* %[[VAL_1118]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1119]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1120:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1120]], float* %[[VAL_26]], float* %[[VAL_1120]]) |
| // CHECK: %[[VAL_1121:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1122:.*]] = getelementptr inbounds float, float* %[[VAL_1121]], i32 %[[VAL_1111]] |
| // CHECK: %[[VAL_1123:.*]] = load float, float* %[[VAL_1122]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1123]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1124:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1124]], float* %[[VAL_24]], float* %[[VAL_1124]]) |
| // CHECK: %[[VAL_1125:.*]] = add i32 1473, %[[VAL_69]] |
| // CHECK: %[[VAL_1126:.*]] = add i32 %[[VAL_70]], 1473 |
| // CHECK: %[[VAL_1127:.*]] = mul nuw nsw i32 %[[VAL_1126]], 1 |
| // CHECK: %[[VAL_1128:.*]] = add nuw nsw i32 0, %[[VAL_1127]] |
| // CHECK: %[[VAL_1129:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1130:.*]] = add nuw nsw i32 %[[VAL_1128]], %[[VAL_1129]] |
| // CHECK: %[[VAL_1131:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1132:.*]] = add nuw nsw i32 %[[VAL_1130]], %[[VAL_1131]] |
| // CHECK: %[[VAL_1133:.*]] = udiv i32 %[[VAL_1132]], 1 |
| // CHECK: %[[VAL_1134:.*]] = urem i32 %[[VAL_1133]], 32 |
| // CHECK: %[[VAL_1135:.*]] = udiv i32 %[[VAL_1132]], 32 |
| // CHECK: %[[VAL_1136:.*]] = urem i32 %[[VAL_1135]], 32 |
| // CHECK: %[[VAL_1137:.*]] = udiv i32 %[[VAL_1132]], 1024 |
| // CHECK: %[[VAL_1138:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1139:.*]] = getelementptr inbounds float, float* %[[VAL_1138]], i32 %[[VAL_1132]] |
| // CHECK: %[[VAL_1140:.*]] = load float, float* %[[VAL_1139]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1140]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1141:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1141]], float* %[[VAL_26]], float* %[[VAL_1141]]) |
| // CHECK: %[[VAL_1142:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1143:.*]] = getelementptr inbounds float, float* %[[VAL_1142]], i32 %[[VAL_1132]] |
| // CHECK: %[[VAL_1144:.*]] = load float, float* %[[VAL_1143]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1144]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1145:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1145]], float* %[[VAL_24]], float* %[[VAL_1145]]) |
| // CHECK: %[[VAL_1146:.*]] = add i32 1536, %[[VAL_69]] |
| // CHECK: %[[VAL_1147:.*]] = add i32 %[[VAL_70]], 1536 |
| // CHECK: %[[VAL_1148:.*]] = mul nuw nsw i32 %[[VAL_1147]], 1 |
| // CHECK: %[[VAL_1149:.*]] = add nuw nsw i32 0, %[[VAL_1148]] |
| // CHECK: %[[VAL_1150:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1151:.*]] = add nuw nsw i32 %[[VAL_1149]], %[[VAL_1150]] |
| // CHECK: %[[VAL_1152:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1153:.*]] = add nuw nsw i32 %[[VAL_1151]], %[[VAL_1152]] |
| // CHECK: %[[VAL_1154:.*]] = udiv i32 %[[VAL_1153]], 1 |
| // CHECK: %[[VAL_1155:.*]] = urem i32 %[[VAL_1154]], 32 |
| // CHECK: %[[VAL_1156:.*]] = udiv i32 %[[VAL_1153]], 32 |
| // CHECK: %[[VAL_1157:.*]] = urem i32 %[[VAL_1156]], 32 |
| // CHECK: %[[VAL_1158:.*]] = udiv i32 %[[VAL_1153]], 1024 |
| // CHECK: %[[VAL_1159:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1160:.*]] = getelementptr inbounds float, float* %[[VAL_1159]], i32 %[[VAL_1153]] |
| // CHECK: %[[VAL_1161:.*]] = load float, float* %[[VAL_1160]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1161]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1162:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1162]], float* %[[VAL_26]], float* %[[VAL_1162]]) |
| // CHECK: %[[VAL_1163:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1164:.*]] = getelementptr inbounds float, float* %[[VAL_1163]], i32 %[[VAL_1153]] |
| // CHECK: %[[VAL_1165:.*]] = load float, float* %[[VAL_1164]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1165]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1166:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1166]], float* %[[VAL_24]], float* %[[VAL_1166]]) |
| // CHECK: %[[VAL_1167:.*]] = add i32 1537, %[[VAL_69]] |
| // CHECK: %[[VAL_1168:.*]] = add i32 %[[VAL_70]], 1537 |
| // CHECK: %[[VAL_1169:.*]] = mul nuw nsw i32 %[[VAL_1168]], 1 |
| // CHECK: %[[VAL_1170:.*]] = add nuw nsw i32 0, %[[VAL_1169]] |
| // CHECK: %[[VAL_1171:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1172:.*]] = add nuw nsw i32 %[[VAL_1170]], %[[VAL_1171]] |
| // CHECK: %[[VAL_1173:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1174:.*]] = add nuw nsw i32 %[[VAL_1172]], %[[VAL_1173]] |
| // CHECK: %[[VAL_1175:.*]] = udiv i32 %[[VAL_1174]], 1 |
| // CHECK: %[[VAL_1176:.*]] = urem i32 %[[VAL_1175]], 32 |
| // CHECK: %[[VAL_1177:.*]] = udiv i32 %[[VAL_1174]], 32 |
| // CHECK: %[[VAL_1178:.*]] = urem i32 %[[VAL_1177]], 32 |
| // CHECK: %[[VAL_1179:.*]] = udiv i32 %[[VAL_1174]], 1024 |
| // CHECK: %[[VAL_1180:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1181:.*]] = getelementptr inbounds float, float* %[[VAL_1180]], i32 %[[VAL_1174]] |
| // CHECK: %[[VAL_1182:.*]] = load float, float* %[[VAL_1181]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1182]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1183:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1183]], float* %[[VAL_26]], float* %[[VAL_1183]]) |
| // CHECK: %[[VAL_1184:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1185:.*]] = getelementptr inbounds float, float* %[[VAL_1184]], i32 %[[VAL_1174]] |
| // CHECK: %[[VAL_1186:.*]] = load float, float* %[[VAL_1185]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1186]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1187:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1187]], float* %[[VAL_24]], float* %[[VAL_1187]]) |
| // CHECK: %[[VAL_1188:.*]] = add i32 1600, %[[VAL_69]] |
| // CHECK: %[[VAL_1189:.*]] = add i32 %[[VAL_70]], 1600 |
| // CHECK: %[[VAL_1190:.*]] = mul nuw nsw i32 %[[VAL_1189]], 1 |
| // CHECK: %[[VAL_1191:.*]] = add nuw nsw i32 0, %[[VAL_1190]] |
| // CHECK: %[[VAL_1192:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1193:.*]] = add nuw nsw i32 %[[VAL_1191]], %[[VAL_1192]] |
| // CHECK: %[[VAL_1194:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1195:.*]] = add nuw nsw i32 %[[VAL_1193]], %[[VAL_1194]] |
| // CHECK: %[[VAL_1196:.*]] = udiv i32 %[[VAL_1195]], 1 |
| // CHECK: %[[VAL_1197:.*]] = urem i32 %[[VAL_1196]], 32 |
| // CHECK: %[[VAL_1198:.*]] = udiv i32 %[[VAL_1195]], 32 |
| // CHECK: %[[VAL_1199:.*]] = urem i32 %[[VAL_1198]], 32 |
| // CHECK: %[[VAL_1200:.*]] = udiv i32 %[[VAL_1195]], 1024 |
| // CHECK: %[[VAL_1201:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1202:.*]] = getelementptr inbounds float, float* %[[VAL_1201]], i32 %[[VAL_1195]] |
| // CHECK: %[[VAL_1203:.*]] = load float, float* %[[VAL_1202]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1203]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1204:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1204]], float* %[[VAL_26]], float* %[[VAL_1204]]) |
| // CHECK: %[[VAL_1205:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1206:.*]] = getelementptr inbounds float, float* %[[VAL_1205]], i32 %[[VAL_1195]] |
| // CHECK: %[[VAL_1207:.*]] = load float, float* %[[VAL_1206]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1207]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1208:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1208]], float* %[[VAL_24]], float* %[[VAL_1208]]) |
| // CHECK: %[[VAL_1209:.*]] = add i32 1601, %[[VAL_69]] |
| // CHECK: %[[VAL_1210:.*]] = add i32 %[[VAL_70]], 1601 |
| // CHECK: %[[VAL_1211:.*]] = mul nuw nsw i32 %[[VAL_1210]], 1 |
| // CHECK: %[[VAL_1212:.*]] = add nuw nsw i32 0, %[[VAL_1211]] |
| // CHECK: %[[VAL_1213:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1214:.*]] = add nuw nsw i32 %[[VAL_1212]], %[[VAL_1213]] |
| // CHECK: %[[VAL_1215:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1216:.*]] = add nuw nsw i32 %[[VAL_1214]], %[[VAL_1215]] |
| // CHECK: %[[VAL_1217:.*]] = udiv i32 %[[VAL_1216]], 1 |
| // CHECK: %[[VAL_1218:.*]] = urem i32 %[[VAL_1217]], 32 |
| // CHECK: %[[VAL_1219:.*]] = udiv i32 %[[VAL_1216]], 32 |
| // CHECK: %[[VAL_1220:.*]] = urem i32 %[[VAL_1219]], 32 |
| // CHECK: %[[VAL_1221:.*]] = udiv i32 %[[VAL_1216]], 1024 |
| // CHECK: %[[VAL_1222:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1223:.*]] = getelementptr inbounds float, float* %[[VAL_1222]], i32 %[[VAL_1216]] |
| // CHECK: %[[VAL_1224:.*]] = load float, float* %[[VAL_1223]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1224]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1225:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1225]], float* %[[VAL_26]], float* %[[VAL_1225]]) |
| // CHECK: %[[VAL_1226:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1227:.*]] = getelementptr inbounds float, float* %[[VAL_1226]], i32 %[[VAL_1216]] |
| // CHECK: %[[VAL_1228:.*]] = load float, float* %[[VAL_1227]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1228]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1229:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1229]], float* %[[VAL_24]], float* %[[VAL_1229]]) |
| // CHECK: %[[VAL_1230:.*]] = add i32 1664, %[[VAL_69]] |
| // CHECK: %[[VAL_1231:.*]] = add i32 %[[VAL_70]], 1664 |
| // CHECK: %[[VAL_1232:.*]] = mul nuw nsw i32 %[[VAL_1231]], 1 |
| // CHECK: %[[VAL_1233:.*]] = add nuw nsw i32 0, %[[VAL_1232]] |
| // CHECK: %[[VAL_1234:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1235:.*]] = add nuw nsw i32 %[[VAL_1233]], %[[VAL_1234]] |
| // CHECK: %[[VAL_1236:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1237:.*]] = add nuw nsw i32 %[[VAL_1235]], %[[VAL_1236]] |
| // CHECK: %[[VAL_1238:.*]] = udiv i32 %[[VAL_1237]], 1 |
| // CHECK: %[[VAL_1239:.*]] = urem i32 %[[VAL_1238]], 32 |
| // CHECK: %[[VAL_1240:.*]] = udiv i32 %[[VAL_1237]], 32 |
| // CHECK: %[[VAL_1241:.*]] = urem i32 %[[VAL_1240]], 32 |
| // CHECK: %[[VAL_1242:.*]] = udiv i32 %[[VAL_1237]], 1024 |
| // CHECK: %[[VAL_1243:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1244:.*]] = getelementptr inbounds float, float* %[[VAL_1243]], i32 %[[VAL_1237]] |
| // CHECK: %[[VAL_1245:.*]] = load float, float* %[[VAL_1244]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1245]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1246:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1246]], float* %[[VAL_26]], float* %[[VAL_1246]]) |
| // CHECK: %[[VAL_1247:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1248:.*]] = getelementptr inbounds float, float* %[[VAL_1247]], i32 %[[VAL_1237]] |
| // CHECK: %[[VAL_1249:.*]] = load float, float* %[[VAL_1248]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1249]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1250:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1250]], float* %[[VAL_24]], float* %[[VAL_1250]]) |
| // CHECK: %[[VAL_1251:.*]] = add i32 1665, %[[VAL_69]] |
| // CHECK: %[[VAL_1252:.*]] = add i32 %[[VAL_70]], 1665 |
| // CHECK: %[[VAL_1253:.*]] = mul nuw nsw i32 %[[VAL_1252]], 1 |
| // CHECK: %[[VAL_1254:.*]] = add nuw nsw i32 0, %[[VAL_1253]] |
| // CHECK: %[[VAL_1255:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1256:.*]] = add nuw nsw i32 %[[VAL_1254]], %[[VAL_1255]] |
| // CHECK: %[[VAL_1257:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1258:.*]] = add nuw nsw i32 %[[VAL_1256]], %[[VAL_1257]] |
| // CHECK: %[[VAL_1259:.*]] = udiv i32 %[[VAL_1258]], 1 |
| // CHECK: %[[VAL_1260:.*]] = urem i32 %[[VAL_1259]], 32 |
| // CHECK: %[[VAL_1261:.*]] = udiv i32 %[[VAL_1258]], 32 |
| // CHECK: %[[VAL_1262:.*]] = urem i32 %[[VAL_1261]], 32 |
| // CHECK: %[[VAL_1263:.*]] = udiv i32 %[[VAL_1258]], 1024 |
| // CHECK: %[[VAL_1264:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1265:.*]] = getelementptr inbounds float, float* %[[VAL_1264]], i32 %[[VAL_1258]] |
| // CHECK: %[[VAL_1266:.*]] = load float, float* %[[VAL_1265]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1266]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1267:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1267]], float* %[[VAL_26]], float* %[[VAL_1267]]) |
| // CHECK: %[[VAL_1268:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1269:.*]] = getelementptr inbounds float, float* %[[VAL_1268]], i32 %[[VAL_1258]] |
| // CHECK: %[[VAL_1270:.*]] = load float, float* %[[VAL_1269]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1270]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1271:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1271]], float* %[[VAL_24]], float* %[[VAL_1271]]) |
| // CHECK: %[[VAL_1272:.*]] = add i32 1728, %[[VAL_69]] |
| // CHECK: %[[VAL_1273:.*]] = add i32 %[[VAL_70]], 1728 |
| // CHECK: %[[VAL_1274:.*]] = mul nuw nsw i32 %[[VAL_1273]], 1 |
| // CHECK: %[[VAL_1275:.*]] = add nuw nsw i32 0, %[[VAL_1274]] |
| // CHECK: %[[VAL_1276:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1277:.*]] = add nuw nsw i32 %[[VAL_1275]], %[[VAL_1276]] |
| // CHECK: %[[VAL_1278:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1279:.*]] = add nuw nsw i32 %[[VAL_1277]], %[[VAL_1278]] |
| // CHECK: %[[VAL_1280:.*]] = udiv i32 %[[VAL_1279]], 1 |
| // CHECK: %[[VAL_1281:.*]] = urem i32 %[[VAL_1280]], 32 |
| // CHECK: %[[VAL_1282:.*]] = udiv i32 %[[VAL_1279]], 32 |
| // CHECK: %[[VAL_1283:.*]] = urem i32 %[[VAL_1282]], 32 |
| // CHECK: %[[VAL_1284:.*]] = udiv i32 %[[VAL_1279]], 1024 |
| // CHECK: %[[VAL_1285:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1286:.*]] = getelementptr inbounds float, float* %[[VAL_1285]], i32 %[[VAL_1279]] |
| // CHECK: %[[VAL_1287:.*]] = load float, float* %[[VAL_1286]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1287]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1288:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1288]], float* %[[VAL_26]], float* %[[VAL_1288]]) |
| // CHECK: %[[VAL_1289:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1290:.*]] = getelementptr inbounds float, float* %[[VAL_1289]], i32 %[[VAL_1279]] |
| // CHECK: %[[VAL_1291:.*]] = load float, float* %[[VAL_1290]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1291]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1292:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1292]], float* %[[VAL_24]], float* %[[VAL_1292]]) |
| // CHECK: %[[VAL_1293:.*]] = add i32 1729, %[[VAL_69]] |
| // CHECK: %[[VAL_1294:.*]] = add i32 %[[VAL_70]], 1729 |
| // CHECK: %[[VAL_1295:.*]] = mul nuw nsw i32 %[[VAL_1294]], 1 |
| // CHECK: %[[VAL_1296:.*]] = add nuw nsw i32 0, %[[VAL_1295]] |
| // CHECK: %[[VAL_1297:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1298:.*]] = add nuw nsw i32 %[[VAL_1296]], %[[VAL_1297]] |
| // CHECK: %[[VAL_1299:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1300:.*]] = add nuw nsw i32 %[[VAL_1298]], %[[VAL_1299]] |
| // CHECK: %[[VAL_1301:.*]] = udiv i32 %[[VAL_1300]], 1 |
| // CHECK: %[[VAL_1302:.*]] = urem i32 %[[VAL_1301]], 32 |
| // CHECK: %[[VAL_1303:.*]] = udiv i32 %[[VAL_1300]], 32 |
| // CHECK: %[[VAL_1304:.*]] = urem i32 %[[VAL_1303]], 32 |
| // CHECK: %[[VAL_1305:.*]] = udiv i32 %[[VAL_1300]], 1024 |
| // CHECK: %[[VAL_1306:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1307:.*]] = getelementptr inbounds float, float* %[[VAL_1306]], i32 %[[VAL_1300]] |
| // CHECK: %[[VAL_1308:.*]] = load float, float* %[[VAL_1307]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1308]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1309:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1309]], float* %[[VAL_26]], float* %[[VAL_1309]]) |
| // CHECK: %[[VAL_1310:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1311:.*]] = getelementptr inbounds float, float* %[[VAL_1310]], i32 %[[VAL_1300]] |
| // CHECK: %[[VAL_1312:.*]] = load float, float* %[[VAL_1311]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1312]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1313:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1313]], float* %[[VAL_24]], float* %[[VAL_1313]]) |
| // CHECK: %[[VAL_1314:.*]] = add i32 1792, %[[VAL_69]] |
| // CHECK: %[[VAL_1315:.*]] = add i32 %[[VAL_70]], 1792 |
| // CHECK: %[[VAL_1316:.*]] = mul nuw nsw i32 %[[VAL_1315]], 1 |
| // CHECK: %[[VAL_1317:.*]] = add nuw nsw i32 0, %[[VAL_1316]] |
| // CHECK: %[[VAL_1318:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1319:.*]] = add nuw nsw i32 %[[VAL_1317]], %[[VAL_1318]] |
| // CHECK: %[[VAL_1320:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1321:.*]] = add nuw nsw i32 %[[VAL_1319]], %[[VAL_1320]] |
| // CHECK: %[[VAL_1322:.*]] = udiv i32 %[[VAL_1321]], 1 |
| // CHECK: %[[VAL_1323:.*]] = urem i32 %[[VAL_1322]], 32 |
| // CHECK: %[[VAL_1324:.*]] = udiv i32 %[[VAL_1321]], 32 |
| // CHECK: %[[VAL_1325:.*]] = urem i32 %[[VAL_1324]], 32 |
| // CHECK: %[[VAL_1326:.*]] = udiv i32 %[[VAL_1321]], 1024 |
| // CHECK: %[[VAL_1327:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1328:.*]] = getelementptr inbounds float, float* %[[VAL_1327]], i32 %[[VAL_1321]] |
| // CHECK: %[[VAL_1329:.*]] = load float, float* %[[VAL_1328]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1329]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1330:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1330]], float* %[[VAL_26]], float* %[[VAL_1330]]) |
| // CHECK: %[[VAL_1331:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1332:.*]] = getelementptr inbounds float, float* %[[VAL_1331]], i32 %[[VAL_1321]] |
| // CHECK: %[[VAL_1333:.*]] = load float, float* %[[VAL_1332]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1333]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1334:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1334]], float* %[[VAL_24]], float* %[[VAL_1334]]) |
| // CHECK: %[[VAL_1335:.*]] = add i32 1793, %[[VAL_69]] |
| // CHECK: %[[VAL_1336:.*]] = add i32 %[[VAL_70]], 1793 |
| // CHECK: %[[VAL_1337:.*]] = mul nuw nsw i32 %[[VAL_1336]], 1 |
| // CHECK: %[[VAL_1338:.*]] = add nuw nsw i32 0, %[[VAL_1337]] |
| // CHECK: %[[VAL_1339:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1340:.*]] = add nuw nsw i32 %[[VAL_1338]], %[[VAL_1339]] |
| // CHECK: %[[VAL_1341:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1342:.*]] = add nuw nsw i32 %[[VAL_1340]], %[[VAL_1341]] |
| // CHECK: %[[VAL_1343:.*]] = udiv i32 %[[VAL_1342]], 1 |
| // CHECK: %[[VAL_1344:.*]] = urem i32 %[[VAL_1343]], 32 |
| // CHECK: %[[VAL_1345:.*]] = udiv i32 %[[VAL_1342]], 32 |
| // CHECK: %[[VAL_1346:.*]] = urem i32 %[[VAL_1345]], 32 |
| // CHECK: %[[VAL_1347:.*]] = udiv i32 %[[VAL_1342]], 1024 |
| // CHECK: %[[VAL_1348:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1349:.*]] = getelementptr inbounds float, float* %[[VAL_1348]], i32 %[[VAL_1342]] |
| // CHECK: %[[VAL_1350:.*]] = load float, float* %[[VAL_1349]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1350]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1351:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1351]], float* %[[VAL_26]], float* %[[VAL_1351]]) |
| // CHECK: %[[VAL_1352:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1353:.*]] = getelementptr inbounds float, float* %[[VAL_1352]], i32 %[[VAL_1342]] |
| // CHECK: %[[VAL_1354:.*]] = load float, float* %[[VAL_1353]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1354]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1355:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1355]], float* %[[VAL_24]], float* %[[VAL_1355]]) |
| // CHECK: %[[VAL_1356:.*]] = add i32 1856, %[[VAL_69]] |
| // CHECK: %[[VAL_1357:.*]] = add i32 %[[VAL_70]], 1856 |
| // CHECK: %[[VAL_1358:.*]] = mul nuw nsw i32 %[[VAL_1357]], 1 |
| // CHECK: %[[VAL_1359:.*]] = add nuw nsw i32 0, %[[VAL_1358]] |
| // CHECK: %[[VAL_1360:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1361:.*]] = add nuw nsw i32 %[[VAL_1359]], %[[VAL_1360]] |
| // CHECK: %[[VAL_1362:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1363:.*]] = add nuw nsw i32 %[[VAL_1361]], %[[VAL_1362]] |
| // CHECK: %[[VAL_1364:.*]] = udiv i32 %[[VAL_1363]], 1 |
| // CHECK: %[[VAL_1365:.*]] = urem i32 %[[VAL_1364]], 32 |
| // CHECK: %[[VAL_1366:.*]] = udiv i32 %[[VAL_1363]], 32 |
| // CHECK: %[[VAL_1367:.*]] = urem i32 %[[VAL_1366]], 32 |
| // CHECK: %[[VAL_1368:.*]] = udiv i32 %[[VAL_1363]], 1024 |
| // CHECK: %[[VAL_1369:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1370:.*]] = getelementptr inbounds float, float* %[[VAL_1369]], i32 %[[VAL_1363]] |
| // CHECK: %[[VAL_1371:.*]] = load float, float* %[[VAL_1370]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1371]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1372:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1372]], float* %[[VAL_26]], float* %[[VAL_1372]]) |
| // CHECK: %[[VAL_1373:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1374:.*]] = getelementptr inbounds float, float* %[[VAL_1373]], i32 %[[VAL_1363]] |
| // CHECK: %[[VAL_1375:.*]] = load float, float* %[[VAL_1374]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1375]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1376:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1376]], float* %[[VAL_24]], float* %[[VAL_1376]]) |
| // CHECK: %[[VAL_1377:.*]] = add i32 1857, %[[VAL_69]] |
| // CHECK: %[[VAL_1378:.*]] = add i32 %[[VAL_70]], 1857 |
| // CHECK: %[[VAL_1379:.*]] = mul nuw nsw i32 %[[VAL_1378]], 1 |
| // CHECK: %[[VAL_1380:.*]] = add nuw nsw i32 0, %[[VAL_1379]] |
| // CHECK: %[[VAL_1381:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1382:.*]] = add nuw nsw i32 %[[VAL_1380]], %[[VAL_1381]] |
| // CHECK: %[[VAL_1383:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1384:.*]] = add nuw nsw i32 %[[VAL_1382]], %[[VAL_1383]] |
| // CHECK: %[[VAL_1385:.*]] = udiv i32 %[[VAL_1384]], 1 |
| // CHECK: %[[VAL_1386:.*]] = urem i32 %[[VAL_1385]], 32 |
| // CHECK: %[[VAL_1387:.*]] = udiv i32 %[[VAL_1384]], 32 |
| // CHECK: %[[VAL_1388:.*]] = urem i32 %[[VAL_1387]], 32 |
| // CHECK: %[[VAL_1389:.*]] = udiv i32 %[[VAL_1384]], 1024 |
| // CHECK: %[[VAL_1390:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1391:.*]] = getelementptr inbounds float, float* %[[VAL_1390]], i32 %[[VAL_1384]] |
| // CHECK: %[[VAL_1392:.*]] = load float, float* %[[VAL_1391]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1392]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1393:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1393]], float* %[[VAL_26]], float* %[[VAL_1393]]) |
| // CHECK: %[[VAL_1394:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1395:.*]] = getelementptr inbounds float, float* %[[VAL_1394]], i32 %[[VAL_1384]] |
| // CHECK: %[[VAL_1396:.*]] = load float, float* %[[VAL_1395]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1396]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1397:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1397]], float* %[[VAL_24]], float* %[[VAL_1397]]) |
| // CHECK: %[[VAL_1398:.*]] = add i32 1920, %[[VAL_69]] |
| // CHECK: %[[VAL_1399:.*]] = add i32 %[[VAL_70]], 1920 |
| // CHECK: %[[VAL_1400:.*]] = mul nuw nsw i32 %[[VAL_1399]], 1 |
| // CHECK: %[[VAL_1401:.*]] = add nuw nsw i32 0, %[[VAL_1400]] |
| // CHECK: %[[VAL_1402:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1403:.*]] = add nuw nsw i32 %[[VAL_1401]], %[[VAL_1402]] |
| // CHECK: %[[VAL_1404:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1405:.*]] = add nuw nsw i32 %[[VAL_1403]], %[[VAL_1404]] |
| // CHECK: %[[VAL_1406:.*]] = udiv i32 %[[VAL_1405]], 1 |
| // CHECK: %[[VAL_1407:.*]] = urem i32 %[[VAL_1406]], 32 |
| // CHECK: %[[VAL_1408:.*]] = udiv i32 %[[VAL_1405]], 32 |
| // CHECK: %[[VAL_1409:.*]] = urem i32 %[[VAL_1408]], 32 |
| // CHECK: %[[VAL_1410:.*]] = udiv i32 %[[VAL_1405]], 1024 |
| // CHECK: %[[VAL_1411:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1412:.*]] = getelementptr inbounds float, float* %[[VAL_1411]], i32 %[[VAL_1405]] |
| // CHECK: %[[VAL_1413:.*]] = load float, float* %[[VAL_1412]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1413]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1414:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1414]], float* %[[VAL_26]], float* %[[VAL_1414]]) |
| // CHECK: %[[VAL_1415:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1416:.*]] = getelementptr inbounds float, float* %[[VAL_1415]], i32 %[[VAL_1405]] |
| // CHECK: %[[VAL_1417:.*]] = load float, float* %[[VAL_1416]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1417]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1418:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1418]], float* %[[VAL_24]], float* %[[VAL_1418]]) |
| // CHECK: %[[VAL_1419:.*]] = add i32 1921, %[[VAL_69]] |
| // CHECK: %[[VAL_1420:.*]] = add i32 %[[VAL_70]], 1921 |
| // CHECK: %[[VAL_1421:.*]] = mul nuw nsw i32 %[[VAL_1420]], 1 |
| // CHECK: %[[VAL_1422:.*]] = add nuw nsw i32 0, %[[VAL_1421]] |
| // CHECK: %[[VAL_1423:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1424:.*]] = add nuw nsw i32 %[[VAL_1422]], %[[VAL_1423]] |
| // CHECK: %[[VAL_1425:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1426:.*]] = add nuw nsw i32 %[[VAL_1424]], %[[VAL_1425]] |
| // CHECK: %[[VAL_1427:.*]] = udiv i32 %[[VAL_1426]], 1 |
| // CHECK: %[[VAL_1428:.*]] = urem i32 %[[VAL_1427]], 32 |
| // CHECK: %[[VAL_1429:.*]] = udiv i32 %[[VAL_1426]], 32 |
| // CHECK: %[[VAL_1430:.*]] = urem i32 %[[VAL_1429]], 32 |
| // CHECK: %[[VAL_1431:.*]] = udiv i32 %[[VAL_1426]], 1024 |
| // CHECK: %[[VAL_1432:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1433:.*]] = getelementptr inbounds float, float* %[[VAL_1432]], i32 %[[VAL_1426]] |
| // CHECK: %[[VAL_1434:.*]] = load float, float* %[[VAL_1433]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1434]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1435:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1435]], float* %[[VAL_26]], float* %[[VAL_1435]]) |
| // CHECK: %[[VAL_1436:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1437:.*]] = getelementptr inbounds float, float* %[[VAL_1436]], i32 %[[VAL_1426]] |
| // CHECK: %[[VAL_1438:.*]] = load float, float* %[[VAL_1437]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1438]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1439:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1439]], float* %[[VAL_24]], float* %[[VAL_1439]]) |
| // CHECK: %[[VAL_1440:.*]] = add i32 1984, %[[VAL_69]] |
| // CHECK: %[[VAL_1441:.*]] = add i32 %[[VAL_70]], 1984 |
| // CHECK: %[[VAL_1442:.*]] = mul nuw nsw i32 %[[VAL_1441]], 1 |
| // CHECK: %[[VAL_1443:.*]] = add nuw nsw i32 0, %[[VAL_1442]] |
| // CHECK: %[[VAL_1444:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1445:.*]] = add nuw nsw i32 %[[VAL_1443]], %[[VAL_1444]] |
| // CHECK: %[[VAL_1446:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1447:.*]] = add nuw nsw i32 %[[VAL_1445]], %[[VAL_1446]] |
| // CHECK: %[[VAL_1448:.*]] = udiv i32 %[[VAL_1447]], 1 |
| // CHECK: %[[VAL_1449:.*]] = urem i32 %[[VAL_1448]], 32 |
| // CHECK: %[[VAL_1450:.*]] = udiv i32 %[[VAL_1447]], 32 |
| // CHECK: %[[VAL_1451:.*]] = urem i32 %[[VAL_1450]], 32 |
| // CHECK: %[[VAL_1452:.*]] = udiv i32 %[[VAL_1447]], 1024 |
| // CHECK: %[[VAL_1453:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1454:.*]] = getelementptr inbounds float, float* %[[VAL_1453]], i32 %[[VAL_1447]] |
| // CHECK: %[[VAL_1455:.*]] = load float, float* %[[VAL_1454]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1455]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1456:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1456]], float* %[[VAL_26]], float* %[[VAL_1456]]) |
| // CHECK: %[[VAL_1457:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1458:.*]] = getelementptr inbounds float, float* %[[VAL_1457]], i32 %[[VAL_1447]] |
| // CHECK: %[[VAL_1459:.*]] = load float, float* %[[VAL_1458]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1459]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1460:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1460]], float* %[[VAL_24]], float* %[[VAL_1460]]) |
| // CHECK: %[[VAL_1461:.*]] = add i32 1985, %[[VAL_69]] |
| // CHECK: %[[VAL_1462:.*]] = add i32 %[[VAL_70]], 1985 |
| // CHECK: %[[VAL_1463:.*]] = mul nuw nsw i32 %[[VAL_1462]], 1 |
| // CHECK: %[[VAL_1464:.*]] = add nuw nsw i32 0, %[[VAL_1463]] |
| // CHECK: %[[VAL_1465:.*]] = mul nuw nsw i32 %[[VAL_137]], 32 |
| // CHECK: %[[VAL_1466:.*]] = add nuw nsw i32 %[[VAL_1464]], %[[VAL_1465]] |
| // CHECK: %[[VAL_1467:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1468:.*]] = add nuw nsw i32 %[[VAL_1466]], %[[VAL_1467]] |
| // CHECK: %[[VAL_1469:.*]] = udiv i32 %[[VAL_1468]], 1 |
| // CHECK: %[[VAL_1470:.*]] = urem i32 %[[VAL_1469]], 32 |
| // CHECK: %[[VAL_1471:.*]] = udiv i32 %[[VAL_1468]], 32 |
| // CHECK: %[[VAL_1472:.*]] = urem i32 %[[VAL_1471]], 32 |
| // CHECK: %[[VAL_1473:.*]] = udiv i32 %[[VAL_1468]], 1024 |
| // CHECK: %[[VAL_1474:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1475:.*]] = getelementptr inbounds float, float* %[[VAL_1474]], i32 %[[VAL_1468]] |
| // CHECK: %[[VAL_1476:.*]] = load float, float* %[[VAL_1475]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1476]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1477:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1477]], float* %[[VAL_26]], float* %[[VAL_1477]]) |
| // CHECK: %[[VAL_1478:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1479:.*]] = getelementptr inbounds float, float* %[[VAL_1478]], i32 %[[VAL_1468]] |
| // CHECK: %[[VAL_1480:.*]] = load float, float* %[[VAL_1479]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1480]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1481:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1481]], float* %[[VAL_24]], float* %[[VAL_1481]]) |
| // CHECK: br label %[[VAL_72]] |
| // CHECK: output_is_full_tile-false: ; preds = %[[VAL_76]] |
| // CHECK: %[[VAL_1482:.*]] = add i32 %[[VAL_67]], %[[VAL_73]] |
| // CHECK: %[[VAL_1483:.*]] = add i32 0, %[[VAL_69]] |
| // CHECK: %[[VAL_1484:.*]] = add i32 %[[VAL_70]], 0 |
| // CHECK: %[[VAL_1485:.*]] = icmp ult i32 %[[VAL_1483]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1485]], label %[[VAL_1486:.*]], label %[[VAL_1487:.*]] |
| // CHECK: output_x_in_tile-after: ; preds = %[[VAL_1486]], %[[VAL_81]] |
| // CHECK: %[[VAL_1488:.*]] = add i32 1, %[[VAL_69]] |
| // CHECK: %[[VAL_1489:.*]] = add i32 %[[VAL_70]], 1 |
| // CHECK: %[[VAL_1490:.*]] = icmp ult i32 %[[VAL_1488]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1490]], label %[[VAL_1491:.*]], label %[[VAL_1492:.*]] |
| // CHECK: output_x_in_tile-after328: ; preds = %[[VAL_1491]], %[[VAL_1487]] |
| // CHECK: %[[VAL_1493:.*]] = add i32 64, %[[VAL_69]] |
| // CHECK: %[[VAL_1494:.*]] = add i32 %[[VAL_70]], 64 |
| // CHECK: %[[VAL_1495:.*]] = icmp ult i32 %[[VAL_1493]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1495]], label %[[VAL_1496:.*]], label %[[VAL_1497:.*]] |
| // CHECK: output_x_in_tile-after335: ; preds = %[[VAL_1496]], %[[VAL_1492]] |
| // CHECK: %[[VAL_1498:.*]] = add i32 65, %[[VAL_69]] |
| // CHECK: %[[VAL_1499:.*]] = add i32 %[[VAL_70]], 65 |
| // CHECK: %[[VAL_1500:.*]] = icmp ult i32 %[[VAL_1498]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1500]], label %[[VAL_1501:.*]], label %[[VAL_1502:.*]] |
| // CHECK: output_x_in_tile-after342: ; preds = %[[VAL_1501]], %[[VAL_1497]] |
| // CHECK: %[[VAL_1503:.*]] = add i32 128, %[[VAL_69]] |
| // CHECK: %[[VAL_1504:.*]] = add i32 %[[VAL_70]], 128 |
| // CHECK: %[[VAL_1505:.*]] = icmp ult i32 %[[VAL_1503]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1505]], label %[[VAL_1506:.*]], label %[[VAL_1507:.*]] |
| // CHECK: output_x_in_tile-after349: ; preds = %[[VAL_1506]], %[[VAL_1502]] |
| // CHECK: %[[VAL_1508:.*]] = add i32 129, %[[VAL_69]] |
| // CHECK: %[[VAL_1509:.*]] = add i32 %[[VAL_70]], 129 |
| // CHECK: %[[VAL_1510:.*]] = icmp ult i32 %[[VAL_1508]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1510]], label %[[VAL_1511:.*]], label %[[VAL_1512:.*]] |
| // CHECK: output_x_in_tile-after356: ; preds = %[[VAL_1511]], %[[VAL_1507]] |
| // CHECK: %[[VAL_1513:.*]] = add i32 192, %[[VAL_69]] |
| // CHECK: %[[VAL_1514:.*]] = add i32 %[[VAL_70]], 192 |
| // CHECK: %[[VAL_1515:.*]] = icmp ult i32 %[[VAL_1513]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1515]], label %[[VAL_1516:.*]], label %[[VAL_1517:.*]] |
| // CHECK: output_x_in_tile-after363: ; preds = %[[VAL_1516]], %[[VAL_1512]] |
| // CHECK: %[[VAL_1518:.*]] = add i32 193, %[[VAL_69]] |
| // CHECK: %[[VAL_1519:.*]] = add i32 %[[VAL_70]], 193 |
| // CHECK: %[[VAL_1520:.*]] = icmp ult i32 %[[VAL_1518]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1520]], label %[[VAL_1521:.*]], label %[[VAL_1522:.*]] |
| // CHECK: output_x_in_tile-after370: ; preds = %[[VAL_1521]], %[[VAL_1517]] |
| // CHECK: %[[VAL_1523:.*]] = add i32 256, %[[VAL_69]] |
| // CHECK: %[[VAL_1524:.*]] = add i32 %[[VAL_70]], 256 |
| // CHECK: %[[VAL_1525:.*]] = icmp ult i32 %[[VAL_1523]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1525]], label %[[VAL_1526:.*]], label %[[VAL_1527:.*]] |
| // CHECK: output_x_in_tile-after377: ; preds = %[[VAL_1526]], %[[VAL_1522]] |
| // CHECK: %[[VAL_1528:.*]] = add i32 257, %[[VAL_69]] |
| // CHECK: %[[VAL_1529:.*]] = add i32 %[[VAL_70]], 257 |
| // CHECK: %[[VAL_1530:.*]] = icmp ult i32 %[[VAL_1528]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1530]], label %[[VAL_1531:.*]], label %[[VAL_1532:.*]] |
| // CHECK: output_x_in_tile-after384: ; preds = %[[VAL_1531]], %[[VAL_1527]] |
| // CHECK: %[[VAL_1533:.*]] = add i32 320, %[[VAL_69]] |
| // CHECK: %[[VAL_1534:.*]] = add i32 %[[VAL_70]], 320 |
| // CHECK: %[[VAL_1535:.*]] = icmp ult i32 %[[VAL_1533]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1535]], label %[[VAL_1536:.*]], label %[[VAL_1537:.*]] |
| // CHECK: output_x_in_tile-after391: ; preds = %[[VAL_1536]], %[[VAL_1532]] |
| // CHECK: %[[VAL_1538:.*]] = add i32 321, %[[VAL_69]] |
| // CHECK: %[[VAL_1539:.*]] = add i32 %[[VAL_70]], 321 |
| // CHECK: %[[VAL_1540:.*]] = icmp ult i32 %[[VAL_1538]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1540]], label %[[VAL_1541:.*]], label %[[VAL_1542:.*]] |
| // CHECK: output_x_in_tile-after398: ; preds = %[[VAL_1541]], %[[VAL_1537]] |
| // CHECK: %[[VAL_1543:.*]] = add i32 384, %[[VAL_69]] |
| // CHECK: %[[VAL_1544:.*]] = add i32 %[[VAL_70]], 384 |
| // CHECK: %[[VAL_1545:.*]] = icmp ult i32 %[[VAL_1543]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1545]], label %[[VAL_1546:.*]], label %[[VAL_1547:.*]] |
| // CHECK: output_x_in_tile-after405: ; preds = %[[VAL_1546]], %[[VAL_1542]] |
| // CHECK: %[[VAL_1548:.*]] = add i32 385, %[[VAL_69]] |
| // CHECK: %[[VAL_1549:.*]] = add i32 %[[VAL_70]], 385 |
| // CHECK: %[[VAL_1550:.*]] = icmp ult i32 %[[VAL_1548]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1550]], label %[[VAL_1551:.*]], label %[[VAL_1552:.*]] |
| // CHECK: output_x_in_tile-after412: ; preds = %[[VAL_1551]], %[[VAL_1547]] |
| // CHECK: %[[VAL_1553:.*]] = add i32 448, %[[VAL_69]] |
| // CHECK: %[[VAL_1554:.*]] = add i32 %[[VAL_70]], 448 |
| // CHECK: %[[VAL_1555:.*]] = icmp ult i32 %[[VAL_1553]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1555]], label %[[VAL_1556:.*]], label %[[VAL_1557:.*]] |
| // CHECK: output_x_in_tile-after419: ; preds = %[[VAL_1556]], %[[VAL_1552]] |
| // CHECK: %[[VAL_1558:.*]] = add i32 449, %[[VAL_69]] |
| // CHECK: %[[VAL_1559:.*]] = add i32 %[[VAL_70]], 449 |
| // CHECK: %[[VAL_1560:.*]] = icmp ult i32 %[[VAL_1558]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1560]], label %[[VAL_1561:.*]], label %[[VAL_1562:.*]] |
| // CHECK: output_x_in_tile-after426: ; preds = %[[VAL_1561]], %[[VAL_1557]] |
| // CHECK: %[[VAL_1563:.*]] = add i32 512, %[[VAL_69]] |
| // CHECK: %[[VAL_1564:.*]] = add i32 %[[VAL_70]], 512 |
| // CHECK: %[[VAL_1565:.*]] = icmp ult i32 %[[VAL_1563]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1565]], label %[[VAL_1566:.*]], label %[[VAL_1567:.*]] |
| // CHECK: output_x_in_tile-after433: ; preds = %[[VAL_1566]], %[[VAL_1562]] |
| // CHECK: %[[VAL_1568:.*]] = add i32 513, %[[VAL_69]] |
| // CHECK: %[[VAL_1569:.*]] = add i32 %[[VAL_70]], 513 |
| // CHECK: %[[VAL_1570:.*]] = icmp ult i32 %[[VAL_1568]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1570]], label %[[VAL_1571:.*]], label %[[VAL_1572:.*]] |
| // CHECK: output_x_in_tile-after440: ; preds = %[[VAL_1571]], %[[VAL_1567]] |
| // CHECK: %[[VAL_1573:.*]] = add i32 576, %[[VAL_69]] |
| // CHECK: %[[VAL_1574:.*]] = add i32 %[[VAL_70]], 576 |
| // CHECK: %[[VAL_1575:.*]] = icmp ult i32 %[[VAL_1573]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1575]], label %[[VAL_1576:.*]], label %[[VAL_1577:.*]] |
| // CHECK: output_x_in_tile-after447: ; preds = %[[VAL_1576]], %[[VAL_1572]] |
| // CHECK: %[[VAL_1578:.*]] = add i32 577, %[[VAL_69]] |
| // CHECK: %[[VAL_1579:.*]] = add i32 %[[VAL_70]], 577 |
| // CHECK: %[[VAL_1580:.*]] = icmp ult i32 %[[VAL_1578]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1580]], label %[[VAL_1581:.*]], label %[[VAL_1582:.*]] |
| // CHECK: output_x_in_tile-after454: ; preds = %[[VAL_1581]], %[[VAL_1577]] |
| // CHECK: %[[VAL_1583:.*]] = add i32 640, %[[VAL_69]] |
| // CHECK: %[[VAL_1584:.*]] = add i32 %[[VAL_70]], 640 |
| // CHECK: %[[VAL_1585:.*]] = icmp ult i32 %[[VAL_1583]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1585]], label %[[VAL_1586:.*]], label %[[VAL_1587:.*]] |
| // CHECK: output_x_in_tile-after461: ; preds = %[[VAL_1586]], %[[VAL_1582]] |
| // CHECK: %[[VAL_1588:.*]] = add i32 641, %[[VAL_69]] |
| // CHECK: %[[VAL_1589:.*]] = add i32 %[[VAL_70]], 641 |
| // CHECK: %[[VAL_1590:.*]] = icmp ult i32 %[[VAL_1588]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1590]], label %[[VAL_1591:.*]], label %[[VAL_1592:.*]] |
| // CHECK: output_x_in_tile-after468: ; preds = %[[VAL_1591]], %[[VAL_1587]] |
| // CHECK: %[[VAL_1593:.*]] = add i32 704, %[[VAL_69]] |
| // CHECK: %[[VAL_1594:.*]] = add i32 %[[VAL_70]], 704 |
| // CHECK: %[[VAL_1595:.*]] = icmp ult i32 %[[VAL_1593]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1595]], label %[[VAL_1596:.*]], label %[[VAL_1597:.*]] |
| // CHECK: output_x_in_tile-after475: ; preds = %[[VAL_1596]], %[[VAL_1592]] |
| // CHECK: %[[VAL_1598:.*]] = add i32 705, %[[VAL_69]] |
| // CHECK: %[[VAL_1599:.*]] = add i32 %[[VAL_70]], 705 |
| // CHECK: %[[VAL_1600:.*]] = icmp ult i32 %[[VAL_1598]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1600]], label %[[VAL_1601:.*]], label %[[VAL_1602:.*]] |
| // CHECK: output_x_in_tile-after482: ; preds = %[[VAL_1601]], %[[VAL_1597]] |
| // CHECK: %[[VAL_1603:.*]] = add i32 768, %[[VAL_69]] |
| // CHECK: %[[VAL_1604:.*]] = add i32 %[[VAL_70]], 768 |
| // CHECK: %[[VAL_1605:.*]] = icmp ult i32 %[[VAL_1603]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1605]], label %[[VAL_1606:.*]], label %[[VAL_1607:.*]] |
| // CHECK: output_x_in_tile-after489: ; preds = %[[VAL_1606]], %[[VAL_1602]] |
| // CHECK: %[[VAL_1608:.*]] = add i32 769, %[[VAL_69]] |
| // CHECK: %[[VAL_1609:.*]] = add i32 %[[VAL_70]], 769 |
| // CHECK: %[[VAL_1610:.*]] = icmp ult i32 %[[VAL_1608]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1610]], label %[[VAL_1611:.*]], label %[[VAL_1612:.*]] |
| // CHECK: output_x_in_tile-after496: ; preds = %[[VAL_1611]], %[[VAL_1607]] |
| // CHECK: %[[VAL_1613:.*]] = add i32 832, %[[VAL_69]] |
| // CHECK: %[[VAL_1614:.*]] = add i32 %[[VAL_70]], 832 |
| // CHECK: %[[VAL_1615:.*]] = icmp ult i32 %[[VAL_1613]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1615]], label %[[VAL_1616:.*]], label %[[VAL_1617:.*]] |
| // CHECK: output_x_in_tile-after503: ; preds = %[[VAL_1616]], %[[VAL_1612]] |
| // CHECK: %[[VAL_1618:.*]] = add i32 833, %[[VAL_69]] |
| // CHECK: %[[VAL_1619:.*]] = add i32 %[[VAL_70]], 833 |
| // CHECK: %[[VAL_1620:.*]] = icmp ult i32 %[[VAL_1618]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1620]], label %[[VAL_1621:.*]], label %[[VAL_1622:.*]] |
| // CHECK: output_x_in_tile-after510: ; preds = %[[VAL_1621]], %[[VAL_1617]] |
| // CHECK: %[[VAL_1623:.*]] = add i32 896, %[[VAL_69]] |
| // CHECK: %[[VAL_1624:.*]] = add i32 %[[VAL_70]], 896 |
| // CHECK: %[[VAL_1625:.*]] = icmp ult i32 %[[VAL_1623]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1625]], label %[[VAL_1626:.*]], label %[[VAL_1627:.*]] |
| // CHECK: output_x_in_tile-after517: ; preds = %[[VAL_1626]], %[[VAL_1622]] |
| // CHECK: %[[VAL_1628:.*]] = add i32 897, %[[VAL_69]] |
| // CHECK: %[[VAL_1629:.*]] = add i32 %[[VAL_70]], 897 |
| // CHECK: %[[VAL_1630:.*]] = icmp ult i32 %[[VAL_1628]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1630]], label %[[VAL_1631:.*]], label %[[VAL_1632:.*]] |
| // CHECK: output_x_in_tile-after524: ; preds = %[[VAL_1631]], %[[VAL_1627]] |
| // CHECK: %[[VAL_1633:.*]] = add i32 960, %[[VAL_69]] |
| // CHECK: %[[VAL_1634:.*]] = add i32 %[[VAL_70]], 960 |
| // CHECK: %[[VAL_1635:.*]] = icmp ult i32 %[[VAL_1633]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1635]], label %[[VAL_1636:.*]], label %[[VAL_1637:.*]] |
| // CHECK: output_x_in_tile-after531: ; preds = %[[VAL_1636]], %[[VAL_1632]] |
| // CHECK: %[[VAL_1638:.*]] = add i32 961, %[[VAL_69]] |
| // CHECK: %[[VAL_1639:.*]] = add i32 %[[VAL_70]], 961 |
| // CHECK: %[[VAL_1640:.*]] = icmp ult i32 %[[VAL_1638]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1640]], label %[[VAL_1641:.*]], label %[[VAL_1642:.*]] |
| // CHECK: output_x_in_tile-after538: ; preds = %[[VAL_1641]], %[[VAL_1637]] |
| // CHECK: %[[VAL_1643:.*]] = add i32 1024, %[[VAL_69]] |
| // CHECK: %[[VAL_1644:.*]] = add i32 %[[VAL_70]], 1024 |
| // CHECK: %[[VAL_1645:.*]] = icmp ult i32 %[[VAL_1643]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1645]], label %[[VAL_1646:.*]], label %[[VAL_1647:.*]] |
| // CHECK: output_x_in_tile-after545: ; preds = %[[VAL_1646]], %[[VAL_1642]] |
| // CHECK: %[[VAL_1648:.*]] = add i32 1025, %[[VAL_69]] |
| // CHECK: %[[VAL_1649:.*]] = add i32 %[[VAL_70]], 1025 |
| // CHECK: %[[VAL_1650:.*]] = icmp ult i32 %[[VAL_1648]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1650]], label %[[VAL_1651:.*]], label %[[VAL_1652:.*]] |
| // CHECK: output_x_in_tile-after552: ; preds = %[[VAL_1651]], %[[VAL_1647]] |
| // CHECK: %[[VAL_1653:.*]] = add i32 1088, %[[VAL_69]] |
| // CHECK: %[[VAL_1654:.*]] = add i32 %[[VAL_70]], 1088 |
| // CHECK: %[[VAL_1655:.*]] = icmp ult i32 %[[VAL_1653]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1655]], label %[[VAL_1656:.*]], label %[[VAL_1657:.*]] |
| // CHECK: output_x_in_tile-after559: ; preds = %[[VAL_1656]], %[[VAL_1652]] |
| // CHECK: %[[VAL_1658:.*]] = add i32 1089, %[[VAL_69]] |
| // CHECK: %[[VAL_1659:.*]] = add i32 %[[VAL_70]], 1089 |
| // CHECK: %[[VAL_1660:.*]] = icmp ult i32 %[[VAL_1658]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1660]], label %[[VAL_1661:.*]], label %[[VAL_1662:.*]] |
| // CHECK: output_x_in_tile-after566: ; preds = %[[VAL_1661]], %[[VAL_1657]] |
| // CHECK: %[[VAL_1663:.*]] = add i32 1152, %[[VAL_69]] |
| // CHECK: %[[VAL_1664:.*]] = add i32 %[[VAL_70]], 1152 |
| // CHECK: %[[VAL_1665:.*]] = icmp ult i32 %[[VAL_1663]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1665]], label %[[VAL_1666:.*]], label %[[VAL_1667:.*]] |
| // CHECK: output_x_in_tile-after573: ; preds = %[[VAL_1666]], %[[VAL_1662]] |
| // CHECK: %[[VAL_1668:.*]] = add i32 1153, %[[VAL_69]] |
| // CHECK: %[[VAL_1669:.*]] = add i32 %[[VAL_70]], 1153 |
| // CHECK: %[[VAL_1670:.*]] = icmp ult i32 %[[VAL_1668]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1670]], label %[[VAL_1671:.*]], label %[[VAL_1672:.*]] |
| // CHECK: output_x_in_tile-after580: ; preds = %[[VAL_1671]], %[[VAL_1667]] |
| // CHECK: %[[VAL_1673:.*]] = add i32 1216, %[[VAL_69]] |
| // CHECK: %[[VAL_1674:.*]] = add i32 %[[VAL_70]], 1216 |
| // CHECK: %[[VAL_1675:.*]] = icmp ult i32 %[[VAL_1673]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1675]], label %[[VAL_1676:.*]], label %[[VAL_1677:.*]] |
| // CHECK: output_x_in_tile-after587: ; preds = %[[VAL_1676]], %[[VAL_1672]] |
| // CHECK: %[[VAL_1678:.*]] = add i32 1217, %[[VAL_69]] |
| // CHECK: %[[VAL_1679:.*]] = add i32 %[[VAL_70]], 1217 |
| // CHECK: %[[VAL_1680:.*]] = icmp ult i32 %[[VAL_1678]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1680]], label %[[VAL_1681:.*]], label %[[VAL_1682:.*]] |
| // CHECK: output_x_in_tile-after594: ; preds = %[[VAL_1681]], %[[VAL_1677]] |
| // CHECK: %[[VAL_1683:.*]] = add i32 1280, %[[VAL_69]] |
| // CHECK: %[[VAL_1684:.*]] = add i32 %[[VAL_70]], 1280 |
| // CHECK: %[[VAL_1685:.*]] = icmp ult i32 %[[VAL_1683]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1685]], label %[[VAL_1686:.*]], label %[[VAL_1687:.*]] |
| // CHECK: output_x_in_tile-after601: ; preds = %[[VAL_1686]], %[[VAL_1682]] |
| // CHECK: %[[VAL_1688:.*]] = add i32 1281, %[[VAL_69]] |
| // CHECK: %[[VAL_1689:.*]] = add i32 %[[VAL_70]], 1281 |
| // CHECK: %[[VAL_1690:.*]] = icmp ult i32 %[[VAL_1688]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1690]], label %[[VAL_1691:.*]], label %[[VAL_1692:.*]] |
| // CHECK: output_x_in_tile-after608: ; preds = %[[VAL_1691]], %[[VAL_1687]] |
| // CHECK: %[[VAL_1693:.*]] = add i32 1344, %[[VAL_69]] |
| // CHECK: %[[VAL_1694:.*]] = add i32 %[[VAL_70]], 1344 |
| // CHECK: %[[VAL_1695:.*]] = icmp ult i32 %[[VAL_1693]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1695]], label %[[VAL_1696:.*]], label %[[VAL_1697:.*]] |
| // CHECK: output_x_in_tile-after615: ; preds = %[[VAL_1696]], %[[VAL_1692]] |
| // CHECK: %[[VAL_1698:.*]] = add i32 1345, %[[VAL_69]] |
| // CHECK: %[[VAL_1699:.*]] = add i32 %[[VAL_70]], 1345 |
| // CHECK: %[[VAL_1700:.*]] = icmp ult i32 %[[VAL_1698]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1700]], label %[[VAL_1701:.*]], label %[[VAL_1702:.*]] |
| // CHECK: output_x_in_tile-after622: ; preds = %[[VAL_1701]], %[[VAL_1697]] |
| // CHECK: %[[VAL_1703:.*]] = add i32 1408, %[[VAL_69]] |
| // CHECK: %[[VAL_1704:.*]] = add i32 %[[VAL_70]], 1408 |
| // CHECK: %[[VAL_1705:.*]] = icmp ult i32 %[[VAL_1703]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1705]], label %[[VAL_1706:.*]], label %[[VAL_1707:.*]] |
| // CHECK: output_x_in_tile-after629: ; preds = %[[VAL_1706]], %[[VAL_1702]] |
| // CHECK: %[[VAL_1708:.*]] = add i32 1409, %[[VAL_69]] |
| // CHECK: %[[VAL_1709:.*]] = add i32 %[[VAL_70]], 1409 |
| // CHECK: %[[VAL_1710:.*]] = icmp ult i32 %[[VAL_1708]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1710]], label %[[VAL_1711:.*]], label %[[VAL_1712:.*]] |
| // CHECK: output_x_in_tile-after636: ; preds = %[[VAL_1711]], %[[VAL_1707]] |
| // CHECK: %[[VAL_1713:.*]] = add i32 1472, %[[VAL_69]] |
| // CHECK: %[[VAL_1714:.*]] = add i32 %[[VAL_70]], 1472 |
| // CHECK: %[[VAL_1715:.*]] = icmp ult i32 %[[VAL_1713]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1715]], label %[[VAL_1716:.*]], label %[[VAL_1717:.*]] |
| // CHECK: output_x_in_tile-after643: ; preds = %[[VAL_1716]], %[[VAL_1712]] |
| // CHECK: %[[VAL_1718:.*]] = add i32 1473, %[[VAL_69]] |
| // CHECK: %[[VAL_1719:.*]] = add i32 %[[VAL_70]], 1473 |
| // CHECK: %[[VAL_1720:.*]] = icmp ult i32 %[[VAL_1718]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1720]], label %[[VAL_1721:.*]], label %[[VAL_1722:.*]] |
| // CHECK: output_x_in_tile-after650: ; preds = %[[VAL_1721]], %[[VAL_1717]] |
| // CHECK: %[[VAL_1723:.*]] = add i32 1536, %[[VAL_69]] |
| // CHECK: %[[VAL_1724:.*]] = add i32 %[[VAL_70]], 1536 |
| // CHECK: %[[VAL_1725:.*]] = icmp ult i32 %[[VAL_1723]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1725]], label %[[VAL_1726:.*]], label %[[VAL_1727:.*]] |
| // CHECK: output_x_in_tile-after657: ; preds = %[[VAL_1726]], %[[VAL_1722]] |
| // CHECK: %[[VAL_1728:.*]] = add i32 1537, %[[VAL_69]] |
| // CHECK: %[[VAL_1729:.*]] = add i32 %[[VAL_70]], 1537 |
| // CHECK: %[[VAL_1730:.*]] = icmp ult i32 %[[VAL_1728]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1730]], label %[[VAL_1731:.*]], label %[[VAL_1732:.*]] |
| // CHECK: output_x_in_tile-after664: ; preds = %[[VAL_1731]], %[[VAL_1727]] |
| // CHECK: %[[VAL_1733:.*]] = add i32 1600, %[[VAL_69]] |
| // CHECK: %[[VAL_1734:.*]] = add i32 %[[VAL_70]], 1600 |
| // CHECK: %[[VAL_1735:.*]] = icmp ult i32 %[[VAL_1733]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1735]], label %[[VAL_1736:.*]], label %[[VAL_1737:.*]] |
| // CHECK: output_x_in_tile-after671: ; preds = %[[VAL_1736]], %[[VAL_1732]] |
| // CHECK: %[[VAL_1738:.*]] = add i32 1601, %[[VAL_69]] |
| // CHECK: %[[VAL_1739:.*]] = add i32 %[[VAL_70]], 1601 |
| // CHECK: %[[VAL_1740:.*]] = icmp ult i32 %[[VAL_1738]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1740]], label %[[VAL_1741:.*]], label %[[VAL_1742:.*]] |
| // CHECK: output_x_in_tile-after678: ; preds = %[[VAL_1741]], %[[VAL_1737]] |
| // CHECK: %[[VAL_1743:.*]] = add i32 1664, %[[VAL_69]] |
| // CHECK: %[[VAL_1744:.*]] = add i32 %[[VAL_70]], 1664 |
| // CHECK: %[[VAL_1745:.*]] = icmp ult i32 %[[VAL_1743]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1745]], label %[[VAL_1746:.*]], label %[[VAL_1747:.*]] |
| // CHECK: output_x_in_tile-after685: ; preds = %[[VAL_1746]], %[[VAL_1742]] |
| // CHECK: %[[VAL_1748:.*]] = add i32 1665, %[[VAL_69]] |
| // CHECK: %[[VAL_1749:.*]] = add i32 %[[VAL_70]], 1665 |
| // CHECK: %[[VAL_1750:.*]] = icmp ult i32 %[[VAL_1748]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1750]], label %[[VAL_1751:.*]], label %[[VAL_1752:.*]] |
| // CHECK: output_x_in_tile-after692: ; preds = %[[VAL_1751]], %[[VAL_1747]] |
| // CHECK: %[[VAL_1753:.*]] = add i32 1728, %[[VAL_69]] |
| // CHECK: %[[VAL_1754:.*]] = add i32 %[[VAL_70]], 1728 |
| // CHECK: %[[VAL_1755:.*]] = icmp ult i32 %[[VAL_1753]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1755]], label %[[VAL_1756:.*]], label %[[VAL_1757:.*]] |
| // CHECK: output_x_in_tile-after699: ; preds = %[[VAL_1756]], %[[VAL_1752]] |
| // CHECK: %[[VAL_1758:.*]] = add i32 1729, %[[VAL_69]] |
| // CHECK: %[[VAL_1759:.*]] = add i32 %[[VAL_70]], 1729 |
| // CHECK: %[[VAL_1760:.*]] = icmp ult i32 %[[VAL_1758]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1760]], label %[[VAL_1761:.*]], label %[[VAL_1762:.*]] |
| // CHECK: output_x_in_tile-after706: ; preds = %[[VAL_1761]], %[[VAL_1757]] |
| // CHECK: %[[VAL_1763:.*]] = add i32 1792, %[[VAL_69]] |
| // CHECK: %[[VAL_1764:.*]] = add i32 %[[VAL_70]], 1792 |
| // CHECK: %[[VAL_1765:.*]] = icmp ult i32 %[[VAL_1763]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1765]], label %[[VAL_1766:.*]], label %[[VAL_1767:.*]] |
| // CHECK: output_x_in_tile-after713: ; preds = %[[VAL_1766]], %[[VAL_1762]] |
| // CHECK: %[[VAL_1768:.*]] = add i32 1793, %[[VAL_69]] |
| // CHECK: %[[VAL_1769:.*]] = add i32 %[[VAL_70]], 1793 |
| // CHECK: %[[VAL_1770:.*]] = icmp ult i32 %[[VAL_1768]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1770]], label %[[VAL_1771:.*]], label %[[VAL_1772:.*]] |
| // CHECK: output_x_in_tile-after720: ; preds = %[[VAL_1771]], %[[VAL_1767]] |
| // CHECK: %[[VAL_1773:.*]] = add i32 1856, %[[VAL_69]] |
| // CHECK: %[[VAL_1774:.*]] = add i32 %[[VAL_70]], 1856 |
| // CHECK: %[[VAL_1775:.*]] = icmp ult i32 %[[VAL_1773]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1775]], label %[[VAL_1776:.*]], label %[[VAL_1777:.*]] |
| // CHECK: output_x_in_tile-after727: ; preds = %[[VAL_1776]], %[[VAL_1772]] |
| // CHECK: %[[VAL_1778:.*]] = add i32 1857, %[[VAL_69]] |
| // CHECK: %[[VAL_1779:.*]] = add i32 %[[VAL_70]], 1857 |
| // CHECK: %[[VAL_1780:.*]] = icmp ult i32 %[[VAL_1778]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1780]], label %[[VAL_1781:.*]], label %[[VAL_1782:.*]] |
| // CHECK: output_x_in_tile-after734: ; preds = %[[VAL_1781]], %[[VAL_1777]] |
| // CHECK: %[[VAL_1783:.*]] = add i32 1920, %[[VAL_69]] |
| // CHECK: %[[VAL_1784:.*]] = add i32 %[[VAL_70]], 1920 |
| // CHECK: %[[VAL_1785:.*]] = icmp ult i32 %[[VAL_1783]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1785]], label %[[VAL_1786:.*]], label %[[VAL_1787:.*]] |
| // CHECK: output_x_in_tile-after741: ; preds = %[[VAL_1786]], %[[VAL_1782]] |
| // CHECK: %[[VAL_1788:.*]] = add i32 1921, %[[VAL_69]] |
| // CHECK: %[[VAL_1789:.*]] = add i32 %[[VAL_70]], 1921 |
| // CHECK: %[[VAL_1790:.*]] = icmp ult i32 %[[VAL_1788]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1790]], label %[[VAL_1791:.*]], label %[[VAL_1792:.*]] |
| // CHECK: output_x_in_tile-after748: ; preds = %[[VAL_1791]], %[[VAL_1787]] |
| // CHECK: %[[VAL_1793:.*]] = add i32 1984, %[[VAL_69]] |
| // CHECK: %[[VAL_1794:.*]] = add i32 %[[VAL_70]], 1984 |
| // CHECK: %[[VAL_1795:.*]] = icmp ult i32 %[[VAL_1793]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1795]], label %[[VAL_1796:.*]], label %[[VAL_1797:.*]] |
| // CHECK: output_x_in_tile-after755: ; preds = %[[VAL_1796]], %[[VAL_1792]] |
| // CHECK: %[[VAL_1798:.*]] = add i32 1985, %[[VAL_69]] |
| // CHECK: %[[VAL_1799:.*]] = add i32 %[[VAL_70]], 1985 |
| // CHECK: %[[VAL_1800:.*]] = icmp ult i32 %[[VAL_1798]], %[[VAL_66]] |
| // CHECK: br i1 %[[VAL_1800]], label %[[VAL_1801:.*]], label %[[VAL_82]] |
| // CHECK: output_x_in_tile-after762: ; preds = %[[VAL_1801]], %[[VAL_1797]] |
| // CHECK: br label %[[VAL_72]] |
| // CHECK: output_x_in_tile-true: ; preds = %[[VAL_81]] |
| // CHECK: %[[VAL_1802:.*]] = mul nuw nsw i32 %[[VAL_1484]], 1 |
| // CHECK: %[[VAL_1803:.*]] = add nuw nsw i32 0, %[[VAL_1802]] |
| // CHECK: %[[VAL_1804:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1805:.*]] = add nuw nsw i32 %[[VAL_1803]], %[[VAL_1804]] |
| // CHECK: %[[VAL_1806:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1807:.*]] = add nuw nsw i32 %[[VAL_1805]], %[[VAL_1806]] |
| // CHECK: %[[VAL_1808:.*]] = udiv i32 %[[VAL_1807]], 1 |
| // CHECK: %[[VAL_1809:.*]] = urem i32 %[[VAL_1808]], 32 |
| // CHECK: %[[VAL_1810:.*]] = udiv i32 %[[VAL_1807]], 32 |
| // CHECK: %[[VAL_1811:.*]] = urem i32 %[[VAL_1810]], 32 |
| // CHECK: %[[VAL_1812:.*]] = udiv i32 %[[VAL_1807]], 1024 |
| // CHECK: %[[VAL_1813:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1814:.*]] = getelementptr inbounds float, float* %[[VAL_1813]], i32 %[[VAL_1807]] |
| // CHECK: %[[VAL_1815:.*]] = load float, float* %[[VAL_1814]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1815]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1816:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1816]], float* %[[VAL_26]], float* %[[VAL_1816]]) |
| // CHECK: %[[VAL_1817:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1818:.*]] = getelementptr inbounds float, float* %[[VAL_1817]], i32 %[[VAL_1807]] |
| // CHECK: %[[VAL_1819:.*]] = load float, float* %[[VAL_1818]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1819]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1820:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1820]], float* %[[VAL_24]], float* %[[VAL_1820]]) |
| // CHECK: br label %[[VAL_1487]] |
| // CHECK: output_x_in_tile-true327: ; preds = %[[VAL_1487]] |
| // CHECK: %[[VAL_1821:.*]] = mul nuw nsw i32 %[[VAL_1489]], 1 |
| // CHECK: %[[VAL_1822:.*]] = add nuw nsw i32 0, %[[VAL_1821]] |
| // CHECK: %[[VAL_1823:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1824:.*]] = add nuw nsw i32 %[[VAL_1822]], %[[VAL_1823]] |
| // CHECK: %[[VAL_1825:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1826:.*]] = add nuw nsw i32 %[[VAL_1824]], %[[VAL_1825]] |
| // CHECK: %[[VAL_1827:.*]] = udiv i32 %[[VAL_1826]], 1 |
| // CHECK: %[[VAL_1828:.*]] = urem i32 %[[VAL_1827]], 32 |
| // CHECK: %[[VAL_1829:.*]] = udiv i32 %[[VAL_1826]], 32 |
| // CHECK: %[[VAL_1830:.*]] = urem i32 %[[VAL_1829]], 32 |
| // CHECK: %[[VAL_1831:.*]] = udiv i32 %[[VAL_1826]], 1024 |
| // CHECK: %[[VAL_1832:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1833:.*]] = getelementptr inbounds float, float* %[[VAL_1832]], i32 %[[VAL_1826]] |
| // CHECK: %[[VAL_1834:.*]] = load float, float* %[[VAL_1833]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1834]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1835:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1835]], float* %[[VAL_26]], float* %[[VAL_1835]]) |
| // CHECK: %[[VAL_1836:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1837:.*]] = getelementptr inbounds float, float* %[[VAL_1836]], i32 %[[VAL_1826]] |
| // CHECK: %[[VAL_1838:.*]] = load float, float* %[[VAL_1837]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1838]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1839:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1839]], float* %[[VAL_24]], float* %[[VAL_1839]]) |
| // CHECK: br label %[[VAL_1492]] |
| // CHECK: output_x_in_tile-true334: ; preds = %[[VAL_1492]] |
| // CHECK: %[[VAL_1840:.*]] = mul nuw nsw i32 %[[VAL_1494]], 1 |
| // CHECK: %[[VAL_1841:.*]] = add nuw nsw i32 0, %[[VAL_1840]] |
| // CHECK: %[[VAL_1842:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1843:.*]] = add nuw nsw i32 %[[VAL_1841]], %[[VAL_1842]] |
| // CHECK: %[[VAL_1844:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1845:.*]] = add nuw nsw i32 %[[VAL_1843]], %[[VAL_1844]] |
| // CHECK: %[[VAL_1846:.*]] = udiv i32 %[[VAL_1845]], 1 |
| // CHECK: %[[VAL_1847:.*]] = urem i32 %[[VAL_1846]], 32 |
| // CHECK: %[[VAL_1848:.*]] = udiv i32 %[[VAL_1845]], 32 |
| // CHECK: %[[VAL_1849:.*]] = urem i32 %[[VAL_1848]], 32 |
| // CHECK: %[[VAL_1850:.*]] = udiv i32 %[[VAL_1845]], 1024 |
| // CHECK: %[[VAL_1851:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1852:.*]] = getelementptr inbounds float, float* %[[VAL_1851]], i32 %[[VAL_1845]] |
| // CHECK: %[[VAL_1853:.*]] = load float, float* %[[VAL_1852]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1853]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1854:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1854]], float* %[[VAL_26]], float* %[[VAL_1854]]) |
| // CHECK: %[[VAL_1855:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1856:.*]] = getelementptr inbounds float, float* %[[VAL_1855]], i32 %[[VAL_1845]] |
| // CHECK: %[[VAL_1857:.*]] = load float, float* %[[VAL_1856]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1857]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1858:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1858]], float* %[[VAL_24]], float* %[[VAL_1858]]) |
| // CHECK: br label %[[VAL_1497]] |
| // CHECK: output_x_in_tile-true341: ; preds = %[[VAL_1497]] |
| // CHECK: %[[VAL_1859:.*]] = mul nuw nsw i32 %[[VAL_1499]], 1 |
| // CHECK: %[[VAL_1860:.*]] = add nuw nsw i32 0, %[[VAL_1859]] |
| // CHECK: %[[VAL_1861:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1862:.*]] = add nuw nsw i32 %[[VAL_1860]], %[[VAL_1861]] |
| // CHECK: %[[VAL_1863:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1864:.*]] = add nuw nsw i32 %[[VAL_1862]], %[[VAL_1863]] |
| // CHECK: %[[VAL_1865:.*]] = udiv i32 %[[VAL_1864]], 1 |
| // CHECK: %[[VAL_1866:.*]] = urem i32 %[[VAL_1865]], 32 |
| // CHECK: %[[VAL_1867:.*]] = udiv i32 %[[VAL_1864]], 32 |
| // CHECK: %[[VAL_1868:.*]] = urem i32 %[[VAL_1867]], 32 |
| // CHECK: %[[VAL_1869:.*]] = udiv i32 %[[VAL_1864]], 1024 |
| // CHECK: %[[VAL_1870:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1871:.*]] = getelementptr inbounds float, float* %[[VAL_1870]], i32 %[[VAL_1864]] |
| // CHECK: %[[VAL_1872:.*]] = load float, float* %[[VAL_1871]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1872]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1873:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1873]], float* %[[VAL_26]], float* %[[VAL_1873]]) |
| // CHECK: %[[VAL_1874:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1875:.*]] = getelementptr inbounds float, float* %[[VAL_1874]], i32 %[[VAL_1864]] |
| // CHECK: %[[VAL_1876:.*]] = load float, float* %[[VAL_1875]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1876]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1877:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1877]], float* %[[VAL_24]], float* %[[VAL_1877]]) |
| // CHECK: br label %[[VAL_1502]] |
| // CHECK: output_x_in_tile-true348: ; preds = %[[VAL_1502]] |
| // CHECK: %[[VAL_1878:.*]] = mul nuw nsw i32 %[[VAL_1504]], 1 |
| // CHECK: %[[VAL_1879:.*]] = add nuw nsw i32 0, %[[VAL_1878]] |
| // CHECK: %[[VAL_1880:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1881:.*]] = add nuw nsw i32 %[[VAL_1879]], %[[VAL_1880]] |
| // CHECK: %[[VAL_1882:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1883:.*]] = add nuw nsw i32 %[[VAL_1881]], %[[VAL_1882]] |
| // CHECK: %[[VAL_1884:.*]] = udiv i32 %[[VAL_1883]], 1 |
| // CHECK: %[[VAL_1885:.*]] = urem i32 %[[VAL_1884]], 32 |
| // CHECK: %[[VAL_1886:.*]] = udiv i32 %[[VAL_1883]], 32 |
| // CHECK: %[[VAL_1887:.*]] = urem i32 %[[VAL_1886]], 32 |
| // CHECK: %[[VAL_1888:.*]] = udiv i32 %[[VAL_1883]], 1024 |
| // CHECK: %[[VAL_1889:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1890:.*]] = getelementptr inbounds float, float* %[[VAL_1889]], i32 %[[VAL_1883]] |
| // CHECK: %[[VAL_1891:.*]] = load float, float* %[[VAL_1890]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1891]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1892:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1892]], float* %[[VAL_26]], float* %[[VAL_1892]]) |
| // CHECK: %[[VAL_1893:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1894:.*]] = getelementptr inbounds float, float* %[[VAL_1893]], i32 %[[VAL_1883]] |
| // CHECK: %[[VAL_1895:.*]] = load float, float* %[[VAL_1894]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1895]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1896:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1896]], float* %[[VAL_24]], float* %[[VAL_1896]]) |
| // CHECK: br label %[[VAL_1507]] |
| // CHECK: output_x_in_tile-true355: ; preds = %[[VAL_1507]] |
| // CHECK: %[[VAL_1897:.*]] = mul nuw nsw i32 %[[VAL_1509]], 1 |
| // CHECK: %[[VAL_1898:.*]] = add nuw nsw i32 0, %[[VAL_1897]] |
| // CHECK: %[[VAL_1899:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1900:.*]] = add nuw nsw i32 %[[VAL_1898]], %[[VAL_1899]] |
| // CHECK: %[[VAL_1901:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1902:.*]] = add nuw nsw i32 %[[VAL_1900]], %[[VAL_1901]] |
| // CHECK: %[[VAL_1903:.*]] = udiv i32 %[[VAL_1902]], 1 |
| // CHECK: %[[VAL_1904:.*]] = urem i32 %[[VAL_1903]], 32 |
| // CHECK: %[[VAL_1905:.*]] = udiv i32 %[[VAL_1902]], 32 |
| // CHECK: %[[VAL_1906:.*]] = urem i32 %[[VAL_1905]], 32 |
| // CHECK: %[[VAL_1907:.*]] = udiv i32 %[[VAL_1902]], 1024 |
| // CHECK: %[[VAL_1908:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1909:.*]] = getelementptr inbounds float, float* %[[VAL_1908]], i32 %[[VAL_1902]] |
| // CHECK: %[[VAL_1910:.*]] = load float, float* %[[VAL_1909]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1910]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1911:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1911]], float* %[[VAL_26]], float* %[[VAL_1911]]) |
| // CHECK: %[[VAL_1912:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1913:.*]] = getelementptr inbounds float, float* %[[VAL_1912]], i32 %[[VAL_1902]] |
| // CHECK: %[[VAL_1914:.*]] = load float, float* %[[VAL_1913]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1914]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1915:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1915]], float* %[[VAL_24]], float* %[[VAL_1915]]) |
| // CHECK: br label %[[VAL_1512]] |
| // CHECK: output_x_in_tile-true362: ; preds = %[[VAL_1512]] |
| // CHECK: %[[VAL_1916:.*]] = mul nuw nsw i32 %[[VAL_1514]], 1 |
| // CHECK: %[[VAL_1917:.*]] = add nuw nsw i32 0, %[[VAL_1916]] |
| // CHECK: %[[VAL_1918:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1919:.*]] = add nuw nsw i32 %[[VAL_1917]], %[[VAL_1918]] |
| // CHECK: %[[VAL_1920:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1921:.*]] = add nuw nsw i32 %[[VAL_1919]], %[[VAL_1920]] |
| // CHECK: %[[VAL_1922:.*]] = udiv i32 %[[VAL_1921]], 1 |
| // CHECK: %[[VAL_1923:.*]] = urem i32 %[[VAL_1922]], 32 |
| // CHECK: %[[VAL_1924:.*]] = udiv i32 %[[VAL_1921]], 32 |
| // CHECK: %[[VAL_1925:.*]] = urem i32 %[[VAL_1924]], 32 |
| // CHECK: %[[VAL_1926:.*]] = udiv i32 %[[VAL_1921]], 1024 |
| // CHECK: %[[VAL_1927:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1928:.*]] = getelementptr inbounds float, float* %[[VAL_1927]], i32 %[[VAL_1921]] |
| // CHECK: %[[VAL_1929:.*]] = load float, float* %[[VAL_1928]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1929]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1930:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1930]], float* %[[VAL_26]], float* %[[VAL_1930]]) |
| // CHECK: %[[VAL_1931:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1932:.*]] = getelementptr inbounds float, float* %[[VAL_1931]], i32 %[[VAL_1921]] |
| // CHECK: %[[VAL_1933:.*]] = load float, float* %[[VAL_1932]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1933]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1934:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1934]], float* %[[VAL_24]], float* %[[VAL_1934]]) |
| // CHECK: br label %[[VAL_1517]] |
| // CHECK: output_x_in_tile-true369: ; preds = %[[VAL_1517]] |
| // CHECK: %[[VAL_1935:.*]] = mul nuw nsw i32 %[[VAL_1519]], 1 |
| // CHECK: %[[VAL_1936:.*]] = add nuw nsw i32 0, %[[VAL_1935]] |
| // CHECK: %[[VAL_1937:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1938:.*]] = add nuw nsw i32 %[[VAL_1936]], %[[VAL_1937]] |
| // CHECK: %[[VAL_1939:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1940:.*]] = add nuw nsw i32 %[[VAL_1938]], %[[VAL_1939]] |
| // CHECK: %[[VAL_1941:.*]] = udiv i32 %[[VAL_1940]], 1 |
| // CHECK: %[[VAL_1942:.*]] = urem i32 %[[VAL_1941]], 32 |
| // CHECK: %[[VAL_1943:.*]] = udiv i32 %[[VAL_1940]], 32 |
| // CHECK: %[[VAL_1944:.*]] = urem i32 %[[VAL_1943]], 32 |
| // CHECK: %[[VAL_1945:.*]] = udiv i32 %[[VAL_1940]], 1024 |
| // CHECK: %[[VAL_1946:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1947:.*]] = getelementptr inbounds float, float* %[[VAL_1946]], i32 %[[VAL_1940]] |
| // CHECK: %[[VAL_1948:.*]] = load float, float* %[[VAL_1947]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1948]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1949:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1949]], float* %[[VAL_26]], float* %[[VAL_1949]]) |
| // CHECK: %[[VAL_1950:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1951:.*]] = getelementptr inbounds float, float* %[[VAL_1950]], i32 %[[VAL_1940]] |
| // CHECK: %[[VAL_1952:.*]] = load float, float* %[[VAL_1951]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1952]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1953:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1953]], float* %[[VAL_24]], float* %[[VAL_1953]]) |
| // CHECK: br label %[[VAL_1522]] |
| // CHECK: output_x_in_tile-true376: ; preds = %[[VAL_1522]] |
| // CHECK: %[[VAL_1954:.*]] = mul nuw nsw i32 %[[VAL_1524]], 1 |
| // CHECK: %[[VAL_1955:.*]] = add nuw nsw i32 0, %[[VAL_1954]] |
| // CHECK: %[[VAL_1956:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1957:.*]] = add nuw nsw i32 %[[VAL_1955]], %[[VAL_1956]] |
| // CHECK: %[[VAL_1958:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1959:.*]] = add nuw nsw i32 %[[VAL_1957]], %[[VAL_1958]] |
| // CHECK: %[[VAL_1960:.*]] = udiv i32 %[[VAL_1959]], 1 |
| // CHECK: %[[VAL_1961:.*]] = urem i32 %[[VAL_1960]], 32 |
| // CHECK: %[[VAL_1962:.*]] = udiv i32 %[[VAL_1959]], 32 |
| // CHECK: %[[VAL_1963:.*]] = urem i32 %[[VAL_1962]], 32 |
| // CHECK: %[[VAL_1964:.*]] = udiv i32 %[[VAL_1959]], 1024 |
| // CHECK: %[[VAL_1965:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1966:.*]] = getelementptr inbounds float, float* %[[VAL_1965]], i32 %[[VAL_1959]] |
| // CHECK: %[[VAL_1967:.*]] = load float, float* %[[VAL_1966]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1967]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1968:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1968]], float* %[[VAL_26]], float* %[[VAL_1968]]) |
| // CHECK: %[[VAL_1969:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1970:.*]] = getelementptr inbounds float, float* %[[VAL_1969]], i32 %[[VAL_1959]] |
| // CHECK: %[[VAL_1971:.*]] = load float, float* %[[VAL_1970]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1971]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1972:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1972]], float* %[[VAL_24]], float* %[[VAL_1972]]) |
| // CHECK: br label %[[VAL_1527]] |
| // CHECK: output_x_in_tile-true383: ; preds = %[[VAL_1527]] |
| // CHECK: %[[VAL_1973:.*]] = mul nuw nsw i32 %[[VAL_1529]], 1 |
| // CHECK: %[[VAL_1974:.*]] = add nuw nsw i32 0, %[[VAL_1973]] |
| // CHECK: %[[VAL_1975:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1976:.*]] = add nuw nsw i32 %[[VAL_1974]], %[[VAL_1975]] |
| // CHECK: %[[VAL_1977:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1978:.*]] = add nuw nsw i32 %[[VAL_1976]], %[[VAL_1977]] |
| // CHECK: %[[VAL_1979:.*]] = udiv i32 %[[VAL_1978]], 1 |
| // CHECK: %[[VAL_1980:.*]] = urem i32 %[[VAL_1979]], 32 |
| // CHECK: %[[VAL_1981:.*]] = udiv i32 %[[VAL_1978]], 32 |
| // CHECK: %[[VAL_1982:.*]] = urem i32 %[[VAL_1981]], 32 |
| // CHECK: %[[VAL_1983:.*]] = udiv i32 %[[VAL_1978]], 1024 |
| // CHECK: %[[VAL_1984:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1985:.*]] = getelementptr inbounds float, float* %[[VAL_1984]], i32 %[[VAL_1978]] |
| // CHECK: %[[VAL_1986:.*]] = load float, float* %[[VAL_1985]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1986]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_1987:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_1987]], float* %[[VAL_26]], float* %[[VAL_1987]]) |
| // CHECK: %[[VAL_1988:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_1989:.*]] = getelementptr inbounds float, float* %[[VAL_1988]], i32 %[[VAL_1978]] |
| // CHECK: %[[VAL_1990:.*]] = load float, float* %[[VAL_1989]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_1990]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_1991:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_1991]], float* %[[VAL_24]], float* %[[VAL_1991]]) |
| // CHECK: br label %[[VAL_1532]] |
| // CHECK: output_x_in_tile-true390: ; preds = %[[VAL_1532]] |
| // CHECK: %[[VAL_1992:.*]] = mul nuw nsw i32 %[[VAL_1534]], 1 |
| // CHECK: %[[VAL_1993:.*]] = add nuw nsw i32 0, %[[VAL_1992]] |
| // CHECK: %[[VAL_1994:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_1995:.*]] = add nuw nsw i32 %[[VAL_1993]], %[[VAL_1994]] |
| // CHECK: %[[VAL_1996:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_1997:.*]] = add nuw nsw i32 %[[VAL_1995]], %[[VAL_1996]] |
| // CHECK: %[[VAL_1998:.*]] = udiv i32 %[[VAL_1997]], 1 |
| // CHECK: %[[VAL_1999:.*]] = urem i32 %[[VAL_1998]], 32 |
| // CHECK: %[[VAL_2000:.*]] = udiv i32 %[[VAL_1997]], 32 |
| // CHECK: %[[VAL_2001:.*]] = urem i32 %[[VAL_2000]], 32 |
| // CHECK: %[[VAL_2002:.*]] = udiv i32 %[[VAL_1997]], 1024 |
| // CHECK: %[[VAL_2003:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2004:.*]] = getelementptr inbounds float, float* %[[VAL_2003]], i32 %[[VAL_1997]] |
| // CHECK: %[[VAL_2005:.*]] = load float, float* %[[VAL_2004]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2005]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2006:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2006]], float* %[[VAL_26]], float* %[[VAL_2006]]) |
| // CHECK: %[[VAL_2007:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2008:.*]] = getelementptr inbounds float, float* %[[VAL_2007]], i32 %[[VAL_1997]] |
| // CHECK: %[[VAL_2009:.*]] = load float, float* %[[VAL_2008]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2009]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2010:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2010]], float* %[[VAL_24]], float* %[[VAL_2010]]) |
| // CHECK: br label %[[VAL_1537]] |
| // CHECK: output_x_in_tile-true397: ; preds = %[[VAL_1537]] |
| // CHECK: %[[VAL_2011:.*]] = mul nuw nsw i32 %[[VAL_1539]], 1 |
| // CHECK: %[[VAL_2012:.*]] = add nuw nsw i32 0, %[[VAL_2011]] |
| // CHECK: %[[VAL_2013:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2014:.*]] = add nuw nsw i32 %[[VAL_2012]], %[[VAL_2013]] |
| // CHECK: %[[VAL_2015:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2016:.*]] = add nuw nsw i32 %[[VAL_2014]], %[[VAL_2015]] |
| // CHECK: %[[VAL_2017:.*]] = udiv i32 %[[VAL_2016]], 1 |
| // CHECK: %[[VAL_2018:.*]] = urem i32 %[[VAL_2017]], 32 |
| // CHECK: %[[VAL_2019:.*]] = udiv i32 %[[VAL_2016]], 32 |
| // CHECK: %[[VAL_2020:.*]] = urem i32 %[[VAL_2019]], 32 |
| // CHECK: %[[VAL_2021:.*]] = udiv i32 %[[VAL_2016]], 1024 |
| // CHECK: %[[VAL_2022:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2023:.*]] = getelementptr inbounds float, float* %[[VAL_2022]], i32 %[[VAL_2016]] |
| // CHECK: %[[VAL_2024:.*]] = load float, float* %[[VAL_2023]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2024]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2025:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2025]], float* %[[VAL_26]], float* %[[VAL_2025]]) |
| // CHECK: %[[VAL_2026:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2027:.*]] = getelementptr inbounds float, float* %[[VAL_2026]], i32 %[[VAL_2016]] |
| // CHECK: %[[VAL_2028:.*]] = load float, float* %[[VAL_2027]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2028]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2029:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2029]], float* %[[VAL_24]], float* %[[VAL_2029]]) |
| // CHECK: br label %[[VAL_1542]] |
| // CHECK: output_x_in_tile-true404: ; preds = %[[VAL_1542]] |
| // CHECK: %[[VAL_2030:.*]] = mul nuw nsw i32 %[[VAL_1544]], 1 |
| // CHECK: %[[VAL_2031:.*]] = add nuw nsw i32 0, %[[VAL_2030]] |
| // CHECK: %[[VAL_2032:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2033:.*]] = add nuw nsw i32 %[[VAL_2031]], %[[VAL_2032]] |
| // CHECK: %[[VAL_2034:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2035:.*]] = add nuw nsw i32 %[[VAL_2033]], %[[VAL_2034]] |
| // CHECK: %[[VAL_2036:.*]] = udiv i32 %[[VAL_2035]], 1 |
| // CHECK: %[[VAL_2037:.*]] = urem i32 %[[VAL_2036]], 32 |
| // CHECK: %[[VAL_2038:.*]] = udiv i32 %[[VAL_2035]], 32 |
| // CHECK: %[[VAL_2039:.*]] = urem i32 %[[VAL_2038]], 32 |
| // CHECK: %[[VAL_2040:.*]] = udiv i32 %[[VAL_2035]], 1024 |
| // CHECK: %[[VAL_2041:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2042:.*]] = getelementptr inbounds float, float* %[[VAL_2041]], i32 %[[VAL_2035]] |
| // CHECK: %[[VAL_2043:.*]] = load float, float* %[[VAL_2042]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2043]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2044:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2044]], float* %[[VAL_26]], float* %[[VAL_2044]]) |
| // CHECK: %[[VAL_2045:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2046:.*]] = getelementptr inbounds float, float* %[[VAL_2045]], i32 %[[VAL_2035]] |
| // CHECK: %[[VAL_2047:.*]] = load float, float* %[[VAL_2046]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2047]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2048:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2048]], float* %[[VAL_24]], float* %[[VAL_2048]]) |
| // CHECK: br label %[[VAL_1547]] |
| // CHECK: output_x_in_tile-true411: ; preds = %[[VAL_1547]] |
| // CHECK: %[[VAL_2049:.*]] = mul nuw nsw i32 %[[VAL_1549]], 1 |
| // CHECK: %[[VAL_2050:.*]] = add nuw nsw i32 0, %[[VAL_2049]] |
| // CHECK: %[[VAL_2051:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2052:.*]] = add nuw nsw i32 %[[VAL_2050]], %[[VAL_2051]] |
| // CHECK: %[[VAL_2053:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2054:.*]] = add nuw nsw i32 %[[VAL_2052]], %[[VAL_2053]] |
| // CHECK: %[[VAL_2055:.*]] = udiv i32 %[[VAL_2054]], 1 |
| // CHECK: %[[VAL_2056:.*]] = urem i32 %[[VAL_2055]], 32 |
| // CHECK: %[[VAL_2057:.*]] = udiv i32 %[[VAL_2054]], 32 |
| // CHECK: %[[VAL_2058:.*]] = urem i32 %[[VAL_2057]], 32 |
| // CHECK: %[[VAL_2059:.*]] = udiv i32 %[[VAL_2054]], 1024 |
| // CHECK: %[[VAL_2060:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2061:.*]] = getelementptr inbounds float, float* %[[VAL_2060]], i32 %[[VAL_2054]] |
| // CHECK: %[[VAL_2062:.*]] = load float, float* %[[VAL_2061]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2062]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2063:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2063]], float* %[[VAL_26]], float* %[[VAL_2063]]) |
| // CHECK: %[[VAL_2064:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2065:.*]] = getelementptr inbounds float, float* %[[VAL_2064]], i32 %[[VAL_2054]] |
| // CHECK: %[[VAL_2066:.*]] = load float, float* %[[VAL_2065]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2066]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2067:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2067]], float* %[[VAL_24]], float* %[[VAL_2067]]) |
| // CHECK: br label %[[VAL_1552]] |
| // CHECK: output_x_in_tile-true418: ; preds = %[[VAL_1552]] |
| // CHECK: %[[VAL_2068:.*]] = mul nuw nsw i32 %[[VAL_1554]], 1 |
| // CHECK: %[[VAL_2069:.*]] = add nuw nsw i32 0, %[[VAL_2068]] |
| // CHECK: %[[VAL_2070:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2071:.*]] = add nuw nsw i32 %[[VAL_2069]], %[[VAL_2070]] |
| // CHECK: %[[VAL_2072:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2073:.*]] = add nuw nsw i32 %[[VAL_2071]], %[[VAL_2072]] |
| // CHECK: %[[VAL_2074:.*]] = udiv i32 %[[VAL_2073]], 1 |
| // CHECK: %[[VAL_2075:.*]] = urem i32 %[[VAL_2074]], 32 |
| // CHECK: %[[VAL_2076:.*]] = udiv i32 %[[VAL_2073]], 32 |
| // CHECK: %[[VAL_2077:.*]] = urem i32 %[[VAL_2076]], 32 |
| // CHECK: %[[VAL_2078:.*]] = udiv i32 %[[VAL_2073]], 1024 |
| // CHECK: %[[VAL_2079:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2080:.*]] = getelementptr inbounds float, float* %[[VAL_2079]], i32 %[[VAL_2073]] |
| // CHECK: %[[VAL_2081:.*]] = load float, float* %[[VAL_2080]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2081]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2082:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2082]], float* %[[VAL_26]], float* %[[VAL_2082]]) |
| // CHECK: %[[VAL_2083:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2084:.*]] = getelementptr inbounds float, float* %[[VAL_2083]], i32 %[[VAL_2073]] |
| // CHECK: %[[VAL_2085:.*]] = load float, float* %[[VAL_2084]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2085]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2086:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2086]], float* %[[VAL_24]], float* %[[VAL_2086]]) |
| // CHECK: br label %[[VAL_1557]] |
| // CHECK: output_x_in_tile-true425: ; preds = %[[VAL_1557]] |
| // CHECK: %[[VAL_2087:.*]] = mul nuw nsw i32 %[[VAL_1559]], 1 |
| // CHECK: %[[VAL_2088:.*]] = add nuw nsw i32 0, %[[VAL_2087]] |
| // CHECK: %[[VAL_2089:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2090:.*]] = add nuw nsw i32 %[[VAL_2088]], %[[VAL_2089]] |
| // CHECK: %[[VAL_2091:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2092:.*]] = add nuw nsw i32 %[[VAL_2090]], %[[VAL_2091]] |
| // CHECK: %[[VAL_2093:.*]] = udiv i32 %[[VAL_2092]], 1 |
| // CHECK: %[[VAL_2094:.*]] = urem i32 %[[VAL_2093]], 32 |
| // CHECK: %[[VAL_2095:.*]] = udiv i32 %[[VAL_2092]], 32 |
| // CHECK: %[[VAL_2096:.*]] = urem i32 %[[VAL_2095]], 32 |
| // CHECK: %[[VAL_2097:.*]] = udiv i32 %[[VAL_2092]], 1024 |
| // CHECK: %[[VAL_2098:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2099:.*]] = getelementptr inbounds float, float* %[[VAL_2098]], i32 %[[VAL_2092]] |
| // CHECK: %[[VAL_2100:.*]] = load float, float* %[[VAL_2099]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2100]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2101:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2101]], float* %[[VAL_26]], float* %[[VAL_2101]]) |
| // CHECK: %[[VAL_2102:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2103:.*]] = getelementptr inbounds float, float* %[[VAL_2102]], i32 %[[VAL_2092]] |
| // CHECK: %[[VAL_2104:.*]] = load float, float* %[[VAL_2103]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2104]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2105:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2105]], float* %[[VAL_24]], float* %[[VAL_2105]]) |
| // CHECK: br label %[[VAL_1562]] |
| // CHECK: output_x_in_tile-true432: ; preds = %[[VAL_1562]] |
| // CHECK: %[[VAL_2106:.*]] = mul nuw nsw i32 %[[VAL_1564]], 1 |
| // CHECK: %[[VAL_2107:.*]] = add nuw nsw i32 0, %[[VAL_2106]] |
| // CHECK: %[[VAL_2108:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2109:.*]] = add nuw nsw i32 %[[VAL_2107]], %[[VAL_2108]] |
| // CHECK: %[[VAL_2110:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2111:.*]] = add nuw nsw i32 %[[VAL_2109]], %[[VAL_2110]] |
| // CHECK: %[[VAL_2112:.*]] = udiv i32 %[[VAL_2111]], 1 |
| // CHECK: %[[VAL_2113:.*]] = urem i32 %[[VAL_2112]], 32 |
| // CHECK: %[[VAL_2114:.*]] = udiv i32 %[[VAL_2111]], 32 |
| // CHECK: %[[VAL_2115:.*]] = urem i32 %[[VAL_2114]], 32 |
| // CHECK: %[[VAL_2116:.*]] = udiv i32 %[[VAL_2111]], 1024 |
| // CHECK: %[[VAL_2117:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2118:.*]] = getelementptr inbounds float, float* %[[VAL_2117]], i32 %[[VAL_2111]] |
| // CHECK: %[[VAL_2119:.*]] = load float, float* %[[VAL_2118]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2119]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2120:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2120]], float* %[[VAL_26]], float* %[[VAL_2120]]) |
| // CHECK: %[[VAL_2121:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2122:.*]] = getelementptr inbounds float, float* %[[VAL_2121]], i32 %[[VAL_2111]] |
| // CHECK: %[[VAL_2123:.*]] = load float, float* %[[VAL_2122]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2123]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2124:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2124]], float* %[[VAL_24]], float* %[[VAL_2124]]) |
| // CHECK: br label %[[VAL_1567]] |
| // CHECK: output_x_in_tile-true439: ; preds = %[[VAL_1567]] |
| // CHECK: %[[VAL_2125:.*]] = mul nuw nsw i32 %[[VAL_1569]], 1 |
| // CHECK: %[[VAL_2126:.*]] = add nuw nsw i32 0, %[[VAL_2125]] |
| // CHECK: %[[VAL_2127:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2128:.*]] = add nuw nsw i32 %[[VAL_2126]], %[[VAL_2127]] |
| // CHECK: %[[VAL_2129:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2130:.*]] = add nuw nsw i32 %[[VAL_2128]], %[[VAL_2129]] |
| // CHECK: %[[VAL_2131:.*]] = udiv i32 %[[VAL_2130]], 1 |
| // CHECK: %[[VAL_2132:.*]] = urem i32 %[[VAL_2131]], 32 |
| // CHECK: %[[VAL_2133:.*]] = udiv i32 %[[VAL_2130]], 32 |
| // CHECK: %[[VAL_2134:.*]] = urem i32 %[[VAL_2133]], 32 |
| // CHECK: %[[VAL_2135:.*]] = udiv i32 %[[VAL_2130]], 1024 |
| // CHECK: %[[VAL_2136:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2137:.*]] = getelementptr inbounds float, float* %[[VAL_2136]], i32 %[[VAL_2130]] |
| // CHECK: %[[VAL_2138:.*]] = load float, float* %[[VAL_2137]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2138]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2139:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2139]], float* %[[VAL_26]], float* %[[VAL_2139]]) |
| // CHECK: %[[VAL_2140:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2141:.*]] = getelementptr inbounds float, float* %[[VAL_2140]], i32 %[[VAL_2130]] |
| // CHECK: %[[VAL_2142:.*]] = load float, float* %[[VAL_2141]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2142]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2143:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2143]], float* %[[VAL_24]], float* %[[VAL_2143]]) |
| // CHECK: br label %[[VAL_1572]] |
| // CHECK: output_x_in_tile-true446: ; preds = %[[VAL_1572]] |
| // CHECK: %[[VAL_2144:.*]] = mul nuw nsw i32 %[[VAL_1574]], 1 |
| // CHECK: %[[VAL_2145:.*]] = add nuw nsw i32 0, %[[VAL_2144]] |
| // CHECK: %[[VAL_2146:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2147:.*]] = add nuw nsw i32 %[[VAL_2145]], %[[VAL_2146]] |
| // CHECK: %[[VAL_2148:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2149:.*]] = add nuw nsw i32 %[[VAL_2147]], %[[VAL_2148]] |
| // CHECK: %[[VAL_2150:.*]] = udiv i32 %[[VAL_2149]], 1 |
| // CHECK: %[[VAL_2151:.*]] = urem i32 %[[VAL_2150]], 32 |
| // CHECK: %[[VAL_2152:.*]] = udiv i32 %[[VAL_2149]], 32 |
| // CHECK: %[[VAL_2153:.*]] = urem i32 %[[VAL_2152]], 32 |
| // CHECK: %[[VAL_2154:.*]] = udiv i32 %[[VAL_2149]], 1024 |
| // CHECK: %[[VAL_2155:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2156:.*]] = getelementptr inbounds float, float* %[[VAL_2155]], i32 %[[VAL_2149]] |
| // CHECK: %[[VAL_2157:.*]] = load float, float* %[[VAL_2156]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2157]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2158:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2158]], float* %[[VAL_26]], float* %[[VAL_2158]]) |
| // CHECK: %[[VAL_2159:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2160:.*]] = getelementptr inbounds float, float* %[[VAL_2159]], i32 %[[VAL_2149]] |
| // CHECK: %[[VAL_2161:.*]] = load float, float* %[[VAL_2160]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2161]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2162:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2162]], float* %[[VAL_24]], float* %[[VAL_2162]]) |
| // CHECK: br label %[[VAL_1577]] |
| // CHECK: output_x_in_tile-true453: ; preds = %[[VAL_1577]] |
| // CHECK: %[[VAL_2163:.*]] = mul nuw nsw i32 %[[VAL_1579]], 1 |
| // CHECK: %[[VAL_2164:.*]] = add nuw nsw i32 0, %[[VAL_2163]] |
| // CHECK: %[[VAL_2165:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2166:.*]] = add nuw nsw i32 %[[VAL_2164]], %[[VAL_2165]] |
| // CHECK: %[[VAL_2167:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2168:.*]] = add nuw nsw i32 %[[VAL_2166]], %[[VAL_2167]] |
| // CHECK: %[[VAL_2169:.*]] = udiv i32 %[[VAL_2168]], 1 |
| // CHECK: %[[VAL_2170:.*]] = urem i32 %[[VAL_2169]], 32 |
| // CHECK: %[[VAL_2171:.*]] = udiv i32 %[[VAL_2168]], 32 |
| // CHECK: %[[VAL_2172:.*]] = urem i32 %[[VAL_2171]], 32 |
| // CHECK: %[[VAL_2173:.*]] = udiv i32 %[[VAL_2168]], 1024 |
| // CHECK: %[[VAL_2174:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2175:.*]] = getelementptr inbounds float, float* %[[VAL_2174]], i32 %[[VAL_2168]] |
| // CHECK: %[[VAL_2176:.*]] = load float, float* %[[VAL_2175]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2176]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2177:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2177]], float* %[[VAL_26]], float* %[[VAL_2177]]) |
| // CHECK: %[[VAL_2178:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2179:.*]] = getelementptr inbounds float, float* %[[VAL_2178]], i32 %[[VAL_2168]] |
| // CHECK: %[[VAL_2180:.*]] = load float, float* %[[VAL_2179]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2180]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2181:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2181]], float* %[[VAL_24]], float* %[[VAL_2181]]) |
| // CHECK: br label %[[VAL_1582]] |
| // CHECK: output_x_in_tile-true460: ; preds = %[[VAL_1582]] |
| // CHECK: %[[VAL_2182:.*]] = mul nuw nsw i32 %[[VAL_1584]], 1 |
| // CHECK: %[[VAL_2183:.*]] = add nuw nsw i32 0, %[[VAL_2182]] |
| // CHECK: %[[VAL_2184:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2185:.*]] = add nuw nsw i32 %[[VAL_2183]], %[[VAL_2184]] |
| // CHECK: %[[VAL_2186:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2187:.*]] = add nuw nsw i32 %[[VAL_2185]], %[[VAL_2186]] |
| // CHECK: %[[VAL_2188:.*]] = udiv i32 %[[VAL_2187]], 1 |
| // CHECK: %[[VAL_2189:.*]] = urem i32 %[[VAL_2188]], 32 |
| // CHECK: %[[VAL_2190:.*]] = udiv i32 %[[VAL_2187]], 32 |
| // CHECK: %[[VAL_2191:.*]] = urem i32 %[[VAL_2190]], 32 |
| // CHECK: %[[VAL_2192:.*]] = udiv i32 %[[VAL_2187]], 1024 |
| // CHECK: %[[VAL_2193:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2194:.*]] = getelementptr inbounds float, float* %[[VAL_2193]], i32 %[[VAL_2187]] |
| // CHECK: %[[VAL_2195:.*]] = load float, float* %[[VAL_2194]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2195]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2196:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2196]], float* %[[VAL_26]], float* %[[VAL_2196]]) |
| // CHECK: %[[VAL_2197:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2198:.*]] = getelementptr inbounds float, float* %[[VAL_2197]], i32 %[[VAL_2187]] |
| // CHECK: %[[VAL_2199:.*]] = load float, float* %[[VAL_2198]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2199]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2200:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2200]], float* %[[VAL_24]], float* %[[VAL_2200]]) |
| // CHECK: br label %[[VAL_1587]] |
| // CHECK: output_x_in_tile-true467: ; preds = %[[VAL_1587]] |
| // CHECK: %[[VAL_2201:.*]] = mul nuw nsw i32 %[[VAL_1589]], 1 |
| // CHECK: %[[VAL_2202:.*]] = add nuw nsw i32 0, %[[VAL_2201]] |
| // CHECK: %[[VAL_2203:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2204:.*]] = add nuw nsw i32 %[[VAL_2202]], %[[VAL_2203]] |
| // CHECK: %[[VAL_2205:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2206:.*]] = add nuw nsw i32 %[[VAL_2204]], %[[VAL_2205]] |
| // CHECK: %[[VAL_2207:.*]] = udiv i32 %[[VAL_2206]], 1 |
| // CHECK: %[[VAL_2208:.*]] = urem i32 %[[VAL_2207]], 32 |
| // CHECK: %[[VAL_2209:.*]] = udiv i32 %[[VAL_2206]], 32 |
| // CHECK: %[[VAL_2210:.*]] = urem i32 %[[VAL_2209]], 32 |
| // CHECK: %[[VAL_2211:.*]] = udiv i32 %[[VAL_2206]], 1024 |
| // CHECK: %[[VAL_2212:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2213:.*]] = getelementptr inbounds float, float* %[[VAL_2212]], i32 %[[VAL_2206]] |
| // CHECK: %[[VAL_2214:.*]] = load float, float* %[[VAL_2213]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2214]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2215:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2215]], float* %[[VAL_26]], float* %[[VAL_2215]]) |
| // CHECK: %[[VAL_2216:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2217:.*]] = getelementptr inbounds float, float* %[[VAL_2216]], i32 %[[VAL_2206]] |
| // CHECK: %[[VAL_2218:.*]] = load float, float* %[[VAL_2217]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2218]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2219:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2219]], float* %[[VAL_24]], float* %[[VAL_2219]]) |
| // CHECK: br label %[[VAL_1592]] |
| // CHECK: output_x_in_tile-true474: ; preds = %[[VAL_1592]] |
| // CHECK: %[[VAL_2220:.*]] = mul nuw nsw i32 %[[VAL_1594]], 1 |
| // CHECK: %[[VAL_2221:.*]] = add nuw nsw i32 0, %[[VAL_2220]] |
| // CHECK: %[[VAL_2222:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2223:.*]] = add nuw nsw i32 %[[VAL_2221]], %[[VAL_2222]] |
| // CHECK: %[[VAL_2224:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2225:.*]] = add nuw nsw i32 %[[VAL_2223]], %[[VAL_2224]] |
| // CHECK: %[[VAL_2226:.*]] = udiv i32 %[[VAL_2225]], 1 |
| // CHECK: %[[VAL_2227:.*]] = urem i32 %[[VAL_2226]], 32 |
| // CHECK: %[[VAL_2228:.*]] = udiv i32 %[[VAL_2225]], 32 |
| // CHECK: %[[VAL_2229:.*]] = urem i32 %[[VAL_2228]], 32 |
| // CHECK: %[[VAL_2230:.*]] = udiv i32 %[[VAL_2225]], 1024 |
| // CHECK: %[[VAL_2231:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2232:.*]] = getelementptr inbounds float, float* %[[VAL_2231]], i32 %[[VAL_2225]] |
| // CHECK: %[[VAL_2233:.*]] = load float, float* %[[VAL_2232]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2233]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2234:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2234]], float* %[[VAL_26]], float* %[[VAL_2234]]) |
| // CHECK: %[[VAL_2235:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2236:.*]] = getelementptr inbounds float, float* %[[VAL_2235]], i32 %[[VAL_2225]] |
| // CHECK: %[[VAL_2237:.*]] = load float, float* %[[VAL_2236]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2237]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2238:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2238]], float* %[[VAL_24]], float* %[[VAL_2238]]) |
| // CHECK: br label %[[VAL_1597]] |
| // CHECK: output_x_in_tile-true481: ; preds = %[[VAL_1597]] |
| // CHECK: %[[VAL_2239:.*]] = mul nuw nsw i32 %[[VAL_1599]], 1 |
| // CHECK: %[[VAL_2240:.*]] = add nuw nsw i32 0, %[[VAL_2239]] |
| // CHECK: %[[VAL_2241:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2242:.*]] = add nuw nsw i32 %[[VAL_2240]], %[[VAL_2241]] |
| // CHECK: %[[VAL_2243:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2244:.*]] = add nuw nsw i32 %[[VAL_2242]], %[[VAL_2243]] |
| // CHECK: %[[VAL_2245:.*]] = udiv i32 %[[VAL_2244]], 1 |
| // CHECK: %[[VAL_2246:.*]] = urem i32 %[[VAL_2245]], 32 |
| // CHECK: %[[VAL_2247:.*]] = udiv i32 %[[VAL_2244]], 32 |
| // CHECK: %[[VAL_2248:.*]] = urem i32 %[[VAL_2247]], 32 |
| // CHECK: %[[VAL_2249:.*]] = udiv i32 %[[VAL_2244]], 1024 |
| // CHECK: %[[VAL_2250:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2251:.*]] = getelementptr inbounds float, float* %[[VAL_2250]], i32 %[[VAL_2244]] |
| // CHECK: %[[VAL_2252:.*]] = load float, float* %[[VAL_2251]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2252]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2253:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2253]], float* %[[VAL_26]], float* %[[VAL_2253]]) |
| // CHECK: %[[VAL_2254:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2255:.*]] = getelementptr inbounds float, float* %[[VAL_2254]], i32 %[[VAL_2244]] |
| // CHECK: %[[VAL_2256:.*]] = load float, float* %[[VAL_2255]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2256]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2257:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2257]], float* %[[VAL_24]], float* %[[VAL_2257]]) |
| // CHECK: br label %[[VAL_1602]] |
| // CHECK: output_x_in_tile-true488: ; preds = %[[VAL_1602]] |
| // CHECK: %[[VAL_2258:.*]] = mul nuw nsw i32 %[[VAL_1604]], 1 |
| // CHECK: %[[VAL_2259:.*]] = add nuw nsw i32 0, %[[VAL_2258]] |
| // CHECK: %[[VAL_2260:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2261:.*]] = add nuw nsw i32 %[[VAL_2259]], %[[VAL_2260]] |
| // CHECK: %[[VAL_2262:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2263:.*]] = add nuw nsw i32 %[[VAL_2261]], %[[VAL_2262]] |
| // CHECK: %[[VAL_2264:.*]] = udiv i32 %[[VAL_2263]], 1 |
| // CHECK: %[[VAL_2265:.*]] = urem i32 %[[VAL_2264]], 32 |
| // CHECK: %[[VAL_2266:.*]] = udiv i32 %[[VAL_2263]], 32 |
| // CHECK: %[[VAL_2267:.*]] = urem i32 %[[VAL_2266]], 32 |
| // CHECK: %[[VAL_2268:.*]] = udiv i32 %[[VAL_2263]], 1024 |
| // CHECK: %[[VAL_2269:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2270:.*]] = getelementptr inbounds float, float* %[[VAL_2269]], i32 %[[VAL_2263]] |
| // CHECK: %[[VAL_2271:.*]] = load float, float* %[[VAL_2270]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2271]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2272:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2272]], float* %[[VAL_26]], float* %[[VAL_2272]]) |
| // CHECK: %[[VAL_2273:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2274:.*]] = getelementptr inbounds float, float* %[[VAL_2273]], i32 %[[VAL_2263]] |
| // CHECK: %[[VAL_2275:.*]] = load float, float* %[[VAL_2274]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2275]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2276:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2276]], float* %[[VAL_24]], float* %[[VAL_2276]]) |
| // CHECK: br label %[[VAL_1607]] |
| // CHECK: output_x_in_tile-true495: ; preds = %[[VAL_1607]] |
| // CHECK: %[[VAL_2277:.*]] = mul nuw nsw i32 %[[VAL_1609]], 1 |
| // CHECK: %[[VAL_2278:.*]] = add nuw nsw i32 0, %[[VAL_2277]] |
| // CHECK: %[[VAL_2279:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2280:.*]] = add nuw nsw i32 %[[VAL_2278]], %[[VAL_2279]] |
| // CHECK: %[[VAL_2281:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2282:.*]] = add nuw nsw i32 %[[VAL_2280]], %[[VAL_2281]] |
| // CHECK: %[[VAL_2283:.*]] = udiv i32 %[[VAL_2282]], 1 |
| // CHECK: %[[VAL_2284:.*]] = urem i32 %[[VAL_2283]], 32 |
| // CHECK: %[[VAL_2285:.*]] = udiv i32 %[[VAL_2282]], 32 |
| // CHECK: %[[VAL_2286:.*]] = urem i32 %[[VAL_2285]], 32 |
| // CHECK: %[[VAL_2287:.*]] = udiv i32 %[[VAL_2282]], 1024 |
| // CHECK: %[[VAL_2288:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2289:.*]] = getelementptr inbounds float, float* %[[VAL_2288]], i32 %[[VAL_2282]] |
| // CHECK: %[[VAL_2290:.*]] = load float, float* %[[VAL_2289]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2290]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2291:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2291]], float* %[[VAL_26]], float* %[[VAL_2291]]) |
| // CHECK: %[[VAL_2292:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2293:.*]] = getelementptr inbounds float, float* %[[VAL_2292]], i32 %[[VAL_2282]] |
| // CHECK: %[[VAL_2294:.*]] = load float, float* %[[VAL_2293]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2294]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2295:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2295]], float* %[[VAL_24]], float* %[[VAL_2295]]) |
| // CHECK: br label %[[VAL_1612]] |
| // CHECK: output_x_in_tile-true502: ; preds = %[[VAL_1612]] |
| // CHECK: %[[VAL_2296:.*]] = mul nuw nsw i32 %[[VAL_1614]], 1 |
| // CHECK: %[[VAL_2297:.*]] = add nuw nsw i32 0, %[[VAL_2296]] |
| // CHECK: %[[VAL_2298:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2299:.*]] = add nuw nsw i32 %[[VAL_2297]], %[[VAL_2298]] |
| // CHECK: %[[VAL_2300:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2301:.*]] = add nuw nsw i32 %[[VAL_2299]], %[[VAL_2300]] |
| // CHECK: %[[VAL_2302:.*]] = udiv i32 %[[VAL_2301]], 1 |
| // CHECK: %[[VAL_2303:.*]] = urem i32 %[[VAL_2302]], 32 |
| // CHECK: %[[VAL_2304:.*]] = udiv i32 %[[VAL_2301]], 32 |
| // CHECK: %[[VAL_2305:.*]] = urem i32 %[[VAL_2304]], 32 |
| // CHECK: %[[VAL_2306:.*]] = udiv i32 %[[VAL_2301]], 1024 |
| // CHECK: %[[VAL_2307:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2308:.*]] = getelementptr inbounds float, float* %[[VAL_2307]], i32 %[[VAL_2301]] |
| // CHECK: %[[VAL_2309:.*]] = load float, float* %[[VAL_2308]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2309]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2310:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2310]], float* %[[VAL_26]], float* %[[VAL_2310]]) |
| // CHECK: %[[VAL_2311:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2312:.*]] = getelementptr inbounds float, float* %[[VAL_2311]], i32 %[[VAL_2301]] |
| // CHECK: %[[VAL_2313:.*]] = load float, float* %[[VAL_2312]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2313]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2314:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2314]], float* %[[VAL_24]], float* %[[VAL_2314]]) |
| // CHECK: br label %[[VAL_1617]] |
| // CHECK: output_x_in_tile-true509: ; preds = %[[VAL_1617]] |
| // CHECK: %[[VAL_2315:.*]] = mul nuw nsw i32 %[[VAL_1619]], 1 |
| // CHECK: %[[VAL_2316:.*]] = add nuw nsw i32 0, %[[VAL_2315]] |
| // CHECK: %[[VAL_2317:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2318:.*]] = add nuw nsw i32 %[[VAL_2316]], %[[VAL_2317]] |
| // CHECK: %[[VAL_2319:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2320:.*]] = add nuw nsw i32 %[[VAL_2318]], %[[VAL_2319]] |
| // CHECK: %[[VAL_2321:.*]] = udiv i32 %[[VAL_2320]], 1 |
| // CHECK: %[[VAL_2322:.*]] = urem i32 %[[VAL_2321]], 32 |
| // CHECK: %[[VAL_2323:.*]] = udiv i32 %[[VAL_2320]], 32 |
| // CHECK: %[[VAL_2324:.*]] = urem i32 %[[VAL_2323]], 32 |
| // CHECK: %[[VAL_2325:.*]] = udiv i32 %[[VAL_2320]], 1024 |
| // CHECK: %[[VAL_2326:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2327:.*]] = getelementptr inbounds float, float* %[[VAL_2326]], i32 %[[VAL_2320]] |
| // CHECK: %[[VAL_2328:.*]] = load float, float* %[[VAL_2327]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2328]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2329:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2329]], float* %[[VAL_26]], float* %[[VAL_2329]]) |
| // CHECK: %[[VAL_2330:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2331:.*]] = getelementptr inbounds float, float* %[[VAL_2330]], i32 %[[VAL_2320]] |
| // CHECK: %[[VAL_2332:.*]] = load float, float* %[[VAL_2331]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2332]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2333:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2333]], float* %[[VAL_24]], float* %[[VAL_2333]]) |
| // CHECK: br label %[[VAL_1622]] |
| // CHECK: output_x_in_tile-true516: ; preds = %[[VAL_1622]] |
| // CHECK: %[[VAL_2334:.*]] = mul nuw nsw i32 %[[VAL_1624]], 1 |
| // CHECK: %[[VAL_2335:.*]] = add nuw nsw i32 0, %[[VAL_2334]] |
| // CHECK: %[[VAL_2336:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2337:.*]] = add nuw nsw i32 %[[VAL_2335]], %[[VAL_2336]] |
| // CHECK: %[[VAL_2338:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2339:.*]] = add nuw nsw i32 %[[VAL_2337]], %[[VAL_2338]] |
| // CHECK: %[[VAL_2340:.*]] = udiv i32 %[[VAL_2339]], 1 |
| // CHECK: %[[VAL_2341:.*]] = urem i32 %[[VAL_2340]], 32 |
| // CHECK: %[[VAL_2342:.*]] = udiv i32 %[[VAL_2339]], 32 |
| // CHECK: %[[VAL_2343:.*]] = urem i32 %[[VAL_2342]], 32 |
| // CHECK: %[[VAL_2344:.*]] = udiv i32 %[[VAL_2339]], 1024 |
| // CHECK: %[[VAL_2345:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2346:.*]] = getelementptr inbounds float, float* %[[VAL_2345]], i32 %[[VAL_2339]] |
| // CHECK: %[[VAL_2347:.*]] = load float, float* %[[VAL_2346]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2347]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2348:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2348]], float* %[[VAL_26]], float* %[[VAL_2348]]) |
| // CHECK: %[[VAL_2349:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2350:.*]] = getelementptr inbounds float, float* %[[VAL_2349]], i32 %[[VAL_2339]] |
| // CHECK: %[[VAL_2351:.*]] = load float, float* %[[VAL_2350]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2351]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2352:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2352]], float* %[[VAL_24]], float* %[[VAL_2352]]) |
| // CHECK: br label %[[VAL_1627]] |
| // CHECK: output_x_in_tile-true523: ; preds = %[[VAL_1627]] |
| // CHECK: %[[VAL_2353:.*]] = mul nuw nsw i32 %[[VAL_1629]], 1 |
| // CHECK: %[[VAL_2354:.*]] = add nuw nsw i32 0, %[[VAL_2353]] |
| // CHECK: %[[VAL_2355:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2356:.*]] = add nuw nsw i32 %[[VAL_2354]], %[[VAL_2355]] |
| // CHECK: %[[VAL_2357:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2358:.*]] = add nuw nsw i32 %[[VAL_2356]], %[[VAL_2357]] |
| // CHECK: %[[VAL_2359:.*]] = udiv i32 %[[VAL_2358]], 1 |
| // CHECK: %[[VAL_2360:.*]] = urem i32 %[[VAL_2359]], 32 |
| // CHECK: %[[VAL_2361:.*]] = udiv i32 %[[VAL_2358]], 32 |
| // CHECK: %[[VAL_2362:.*]] = urem i32 %[[VAL_2361]], 32 |
| // CHECK: %[[VAL_2363:.*]] = udiv i32 %[[VAL_2358]], 1024 |
| // CHECK: %[[VAL_2364:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2365:.*]] = getelementptr inbounds float, float* %[[VAL_2364]], i32 %[[VAL_2358]] |
| // CHECK: %[[VAL_2366:.*]] = load float, float* %[[VAL_2365]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2366]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2367:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2367]], float* %[[VAL_26]], float* %[[VAL_2367]]) |
| // CHECK: %[[VAL_2368:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2369:.*]] = getelementptr inbounds float, float* %[[VAL_2368]], i32 %[[VAL_2358]] |
| // CHECK: %[[VAL_2370:.*]] = load float, float* %[[VAL_2369]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2370]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2371:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2371]], float* %[[VAL_24]], float* %[[VAL_2371]]) |
| // CHECK: br label %[[VAL_1632]] |
| // CHECK: output_x_in_tile-true530: ; preds = %[[VAL_1632]] |
| // CHECK: %[[VAL_2372:.*]] = mul nuw nsw i32 %[[VAL_1634]], 1 |
| // CHECK: %[[VAL_2373:.*]] = add nuw nsw i32 0, %[[VAL_2372]] |
| // CHECK: %[[VAL_2374:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2375:.*]] = add nuw nsw i32 %[[VAL_2373]], %[[VAL_2374]] |
| // CHECK: %[[VAL_2376:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2377:.*]] = add nuw nsw i32 %[[VAL_2375]], %[[VAL_2376]] |
| // CHECK: %[[VAL_2378:.*]] = udiv i32 %[[VAL_2377]], 1 |
| // CHECK: %[[VAL_2379:.*]] = urem i32 %[[VAL_2378]], 32 |
| // CHECK: %[[VAL_2380:.*]] = udiv i32 %[[VAL_2377]], 32 |
| // CHECK: %[[VAL_2381:.*]] = urem i32 %[[VAL_2380]], 32 |
| // CHECK: %[[VAL_2382:.*]] = udiv i32 %[[VAL_2377]], 1024 |
| // CHECK: %[[VAL_2383:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2384:.*]] = getelementptr inbounds float, float* %[[VAL_2383]], i32 %[[VAL_2377]] |
| // CHECK: %[[VAL_2385:.*]] = load float, float* %[[VAL_2384]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2385]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2386:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2386]], float* %[[VAL_26]], float* %[[VAL_2386]]) |
| // CHECK: %[[VAL_2387:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2388:.*]] = getelementptr inbounds float, float* %[[VAL_2387]], i32 %[[VAL_2377]] |
| // CHECK: %[[VAL_2389:.*]] = load float, float* %[[VAL_2388]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2389]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2390:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2390]], float* %[[VAL_24]], float* %[[VAL_2390]]) |
| // CHECK: br label %[[VAL_1637]] |
| // CHECK: output_x_in_tile-true537: ; preds = %[[VAL_1637]] |
| // CHECK: %[[VAL_2391:.*]] = mul nuw nsw i32 %[[VAL_1639]], 1 |
| // CHECK: %[[VAL_2392:.*]] = add nuw nsw i32 0, %[[VAL_2391]] |
| // CHECK: %[[VAL_2393:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2394:.*]] = add nuw nsw i32 %[[VAL_2392]], %[[VAL_2393]] |
| // CHECK: %[[VAL_2395:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2396:.*]] = add nuw nsw i32 %[[VAL_2394]], %[[VAL_2395]] |
| // CHECK: %[[VAL_2397:.*]] = udiv i32 %[[VAL_2396]], 1 |
| // CHECK: %[[VAL_2398:.*]] = urem i32 %[[VAL_2397]], 32 |
| // CHECK: %[[VAL_2399:.*]] = udiv i32 %[[VAL_2396]], 32 |
| // CHECK: %[[VAL_2400:.*]] = urem i32 %[[VAL_2399]], 32 |
| // CHECK: %[[VAL_2401:.*]] = udiv i32 %[[VAL_2396]], 1024 |
| // CHECK: %[[VAL_2402:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2403:.*]] = getelementptr inbounds float, float* %[[VAL_2402]], i32 %[[VAL_2396]] |
| // CHECK: %[[VAL_2404:.*]] = load float, float* %[[VAL_2403]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2404]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2405:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2405]], float* %[[VAL_26]], float* %[[VAL_2405]]) |
| // CHECK: %[[VAL_2406:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2407:.*]] = getelementptr inbounds float, float* %[[VAL_2406]], i32 %[[VAL_2396]] |
| // CHECK: %[[VAL_2408:.*]] = load float, float* %[[VAL_2407]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2408]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2409:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2409]], float* %[[VAL_24]], float* %[[VAL_2409]]) |
| // CHECK: br label %[[VAL_1642]] |
| // CHECK: output_x_in_tile-true544: ; preds = %[[VAL_1642]] |
| // CHECK: %[[VAL_2410:.*]] = mul nuw nsw i32 %[[VAL_1644]], 1 |
| // CHECK: %[[VAL_2411:.*]] = add nuw nsw i32 0, %[[VAL_2410]] |
| // CHECK: %[[VAL_2412:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2413:.*]] = add nuw nsw i32 %[[VAL_2411]], %[[VAL_2412]] |
| // CHECK: %[[VAL_2414:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2415:.*]] = add nuw nsw i32 %[[VAL_2413]], %[[VAL_2414]] |
| // CHECK: %[[VAL_2416:.*]] = udiv i32 %[[VAL_2415]], 1 |
| // CHECK: %[[VAL_2417:.*]] = urem i32 %[[VAL_2416]], 32 |
| // CHECK: %[[VAL_2418:.*]] = udiv i32 %[[VAL_2415]], 32 |
| // CHECK: %[[VAL_2419:.*]] = urem i32 %[[VAL_2418]], 32 |
| // CHECK: %[[VAL_2420:.*]] = udiv i32 %[[VAL_2415]], 1024 |
| // CHECK: %[[VAL_2421:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2422:.*]] = getelementptr inbounds float, float* %[[VAL_2421]], i32 %[[VAL_2415]] |
| // CHECK: %[[VAL_2423:.*]] = load float, float* %[[VAL_2422]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2423]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2424:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2424]], float* %[[VAL_26]], float* %[[VAL_2424]]) |
| // CHECK: %[[VAL_2425:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2426:.*]] = getelementptr inbounds float, float* %[[VAL_2425]], i32 %[[VAL_2415]] |
| // CHECK: %[[VAL_2427:.*]] = load float, float* %[[VAL_2426]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2427]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2428:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2428]], float* %[[VAL_24]], float* %[[VAL_2428]]) |
| // CHECK: br label %[[VAL_1647]] |
| // CHECK: output_x_in_tile-true551: ; preds = %[[VAL_1647]] |
| // CHECK: %[[VAL_2429:.*]] = mul nuw nsw i32 %[[VAL_1649]], 1 |
| // CHECK: %[[VAL_2430:.*]] = add nuw nsw i32 0, %[[VAL_2429]] |
| // CHECK: %[[VAL_2431:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2432:.*]] = add nuw nsw i32 %[[VAL_2430]], %[[VAL_2431]] |
| // CHECK: %[[VAL_2433:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2434:.*]] = add nuw nsw i32 %[[VAL_2432]], %[[VAL_2433]] |
| // CHECK: %[[VAL_2435:.*]] = udiv i32 %[[VAL_2434]], 1 |
| // CHECK: %[[VAL_2436:.*]] = urem i32 %[[VAL_2435]], 32 |
| // CHECK: %[[VAL_2437:.*]] = udiv i32 %[[VAL_2434]], 32 |
| // CHECK: %[[VAL_2438:.*]] = urem i32 %[[VAL_2437]], 32 |
| // CHECK: %[[VAL_2439:.*]] = udiv i32 %[[VAL_2434]], 1024 |
| // CHECK: %[[VAL_2440:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2441:.*]] = getelementptr inbounds float, float* %[[VAL_2440]], i32 %[[VAL_2434]] |
| // CHECK: %[[VAL_2442:.*]] = load float, float* %[[VAL_2441]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2442]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2443:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2443]], float* %[[VAL_26]], float* %[[VAL_2443]]) |
| // CHECK: %[[VAL_2444:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2445:.*]] = getelementptr inbounds float, float* %[[VAL_2444]], i32 %[[VAL_2434]] |
| // CHECK: %[[VAL_2446:.*]] = load float, float* %[[VAL_2445]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2446]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2447:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2447]], float* %[[VAL_24]], float* %[[VAL_2447]]) |
| // CHECK: br label %[[VAL_1652]] |
| // CHECK: output_x_in_tile-true558: ; preds = %[[VAL_1652]] |
| // CHECK: %[[VAL_2448:.*]] = mul nuw nsw i32 %[[VAL_1654]], 1 |
| // CHECK: %[[VAL_2449:.*]] = add nuw nsw i32 0, %[[VAL_2448]] |
| // CHECK: %[[VAL_2450:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2451:.*]] = add nuw nsw i32 %[[VAL_2449]], %[[VAL_2450]] |
| // CHECK: %[[VAL_2452:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2453:.*]] = add nuw nsw i32 %[[VAL_2451]], %[[VAL_2452]] |
| // CHECK: %[[VAL_2454:.*]] = udiv i32 %[[VAL_2453]], 1 |
| // CHECK: %[[VAL_2455:.*]] = urem i32 %[[VAL_2454]], 32 |
| // CHECK: %[[VAL_2456:.*]] = udiv i32 %[[VAL_2453]], 32 |
| // CHECK: %[[VAL_2457:.*]] = urem i32 %[[VAL_2456]], 32 |
| // CHECK: %[[VAL_2458:.*]] = udiv i32 %[[VAL_2453]], 1024 |
| // CHECK: %[[VAL_2459:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2460:.*]] = getelementptr inbounds float, float* %[[VAL_2459]], i32 %[[VAL_2453]] |
| // CHECK: %[[VAL_2461:.*]] = load float, float* %[[VAL_2460]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2461]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2462:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2462]], float* %[[VAL_26]], float* %[[VAL_2462]]) |
| // CHECK: %[[VAL_2463:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2464:.*]] = getelementptr inbounds float, float* %[[VAL_2463]], i32 %[[VAL_2453]] |
| // CHECK: %[[VAL_2465:.*]] = load float, float* %[[VAL_2464]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2465]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2466:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2466]], float* %[[VAL_24]], float* %[[VAL_2466]]) |
| // CHECK: br label %[[VAL_1657]] |
| // CHECK: output_x_in_tile-true565: ; preds = %[[VAL_1657]] |
| // CHECK: %[[VAL_2467:.*]] = mul nuw nsw i32 %[[VAL_1659]], 1 |
| // CHECK: %[[VAL_2468:.*]] = add nuw nsw i32 0, %[[VAL_2467]] |
| // CHECK: %[[VAL_2469:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2470:.*]] = add nuw nsw i32 %[[VAL_2468]], %[[VAL_2469]] |
| // CHECK: %[[VAL_2471:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2472:.*]] = add nuw nsw i32 %[[VAL_2470]], %[[VAL_2471]] |
| // CHECK: %[[VAL_2473:.*]] = udiv i32 %[[VAL_2472]], 1 |
| // CHECK: %[[VAL_2474:.*]] = urem i32 %[[VAL_2473]], 32 |
| // CHECK: %[[VAL_2475:.*]] = udiv i32 %[[VAL_2472]], 32 |
| // CHECK: %[[VAL_2476:.*]] = urem i32 %[[VAL_2475]], 32 |
| // CHECK: %[[VAL_2477:.*]] = udiv i32 %[[VAL_2472]], 1024 |
| // CHECK: %[[VAL_2478:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2479:.*]] = getelementptr inbounds float, float* %[[VAL_2478]], i32 %[[VAL_2472]] |
| // CHECK: %[[VAL_2480:.*]] = load float, float* %[[VAL_2479]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2480]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2481:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2481]], float* %[[VAL_26]], float* %[[VAL_2481]]) |
| // CHECK: %[[VAL_2482:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2483:.*]] = getelementptr inbounds float, float* %[[VAL_2482]], i32 %[[VAL_2472]] |
| // CHECK: %[[VAL_2484:.*]] = load float, float* %[[VAL_2483]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2484]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2485:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2485]], float* %[[VAL_24]], float* %[[VAL_2485]]) |
| // CHECK: br label %[[VAL_1662]] |
| // CHECK: output_x_in_tile-true572: ; preds = %[[VAL_1662]] |
| // CHECK: %[[VAL_2486:.*]] = mul nuw nsw i32 %[[VAL_1664]], 1 |
| // CHECK: %[[VAL_2487:.*]] = add nuw nsw i32 0, %[[VAL_2486]] |
| // CHECK: %[[VAL_2488:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2489:.*]] = add nuw nsw i32 %[[VAL_2487]], %[[VAL_2488]] |
| // CHECK: %[[VAL_2490:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2491:.*]] = add nuw nsw i32 %[[VAL_2489]], %[[VAL_2490]] |
| // CHECK: %[[VAL_2492:.*]] = udiv i32 %[[VAL_2491]], 1 |
| // CHECK: %[[VAL_2493:.*]] = urem i32 %[[VAL_2492]], 32 |
| // CHECK: %[[VAL_2494:.*]] = udiv i32 %[[VAL_2491]], 32 |
| // CHECK: %[[VAL_2495:.*]] = urem i32 %[[VAL_2494]], 32 |
| // CHECK: %[[VAL_2496:.*]] = udiv i32 %[[VAL_2491]], 1024 |
| // CHECK: %[[VAL_2497:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2498:.*]] = getelementptr inbounds float, float* %[[VAL_2497]], i32 %[[VAL_2491]] |
| // CHECK: %[[VAL_2499:.*]] = load float, float* %[[VAL_2498]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2499]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2500:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2500]], float* %[[VAL_26]], float* %[[VAL_2500]]) |
| // CHECK: %[[VAL_2501:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2502:.*]] = getelementptr inbounds float, float* %[[VAL_2501]], i32 %[[VAL_2491]] |
| // CHECK: %[[VAL_2503:.*]] = load float, float* %[[VAL_2502]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2503]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2504:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2504]], float* %[[VAL_24]], float* %[[VAL_2504]]) |
| // CHECK: br label %[[VAL_1667]] |
| // CHECK: output_x_in_tile-true579: ; preds = %[[VAL_1667]] |
| // CHECK: %[[VAL_2505:.*]] = mul nuw nsw i32 %[[VAL_1669]], 1 |
| // CHECK: %[[VAL_2506:.*]] = add nuw nsw i32 0, %[[VAL_2505]] |
| // CHECK: %[[VAL_2507:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2508:.*]] = add nuw nsw i32 %[[VAL_2506]], %[[VAL_2507]] |
| // CHECK: %[[VAL_2509:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2510:.*]] = add nuw nsw i32 %[[VAL_2508]], %[[VAL_2509]] |
| // CHECK: %[[VAL_2511:.*]] = udiv i32 %[[VAL_2510]], 1 |
| // CHECK: %[[VAL_2512:.*]] = urem i32 %[[VAL_2511]], 32 |
| // CHECK: %[[VAL_2513:.*]] = udiv i32 %[[VAL_2510]], 32 |
| // CHECK: %[[VAL_2514:.*]] = urem i32 %[[VAL_2513]], 32 |
| // CHECK: %[[VAL_2515:.*]] = udiv i32 %[[VAL_2510]], 1024 |
| // CHECK: %[[VAL_2516:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2517:.*]] = getelementptr inbounds float, float* %[[VAL_2516]], i32 %[[VAL_2510]] |
| // CHECK: %[[VAL_2518:.*]] = load float, float* %[[VAL_2517]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2518]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2519:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2519]], float* %[[VAL_26]], float* %[[VAL_2519]]) |
| // CHECK: %[[VAL_2520:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2521:.*]] = getelementptr inbounds float, float* %[[VAL_2520]], i32 %[[VAL_2510]] |
| // CHECK: %[[VAL_2522:.*]] = load float, float* %[[VAL_2521]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2522]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2523:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2523]], float* %[[VAL_24]], float* %[[VAL_2523]]) |
| // CHECK: br label %[[VAL_1672]] |
| // CHECK: output_x_in_tile-true586: ; preds = %[[VAL_1672]] |
| // CHECK: %[[VAL_2524:.*]] = mul nuw nsw i32 %[[VAL_1674]], 1 |
| // CHECK: %[[VAL_2525:.*]] = add nuw nsw i32 0, %[[VAL_2524]] |
| // CHECK: %[[VAL_2526:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2527:.*]] = add nuw nsw i32 %[[VAL_2525]], %[[VAL_2526]] |
| // CHECK: %[[VAL_2528:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2529:.*]] = add nuw nsw i32 %[[VAL_2527]], %[[VAL_2528]] |
| // CHECK: %[[VAL_2530:.*]] = udiv i32 %[[VAL_2529]], 1 |
| // CHECK: %[[VAL_2531:.*]] = urem i32 %[[VAL_2530]], 32 |
| // CHECK: %[[VAL_2532:.*]] = udiv i32 %[[VAL_2529]], 32 |
| // CHECK: %[[VAL_2533:.*]] = urem i32 %[[VAL_2532]], 32 |
| // CHECK: %[[VAL_2534:.*]] = udiv i32 %[[VAL_2529]], 1024 |
| // CHECK: %[[VAL_2535:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2536:.*]] = getelementptr inbounds float, float* %[[VAL_2535]], i32 %[[VAL_2529]] |
| // CHECK: %[[VAL_2537:.*]] = load float, float* %[[VAL_2536]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2537]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2538:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2538]], float* %[[VAL_26]], float* %[[VAL_2538]]) |
| // CHECK: %[[VAL_2539:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2540:.*]] = getelementptr inbounds float, float* %[[VAL_2539]], i32 %[[VAL_2529]] |
| // CHECK: %[[VAL_2541:.*]] = load float, float* %[[VAL_2540]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2541]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2542:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2542]], float* %[[VAL_24]], float* %[[VAL_2542]]) |
| // CHECK: br label %[[VAL_1677]] |
| // CHECK: output_x_in_tile-true593: ; preds = %[[VAL_1677]] |
| // CHECK: %[[VAL_2543:.*]] = mul nuw nsw i32 %[[VAL_1679]], 1 |
| // CHECK: %[[VAL_2544:.*]] = add nuw nsw i32 0, %[[VAL_2543]] |
| // CHECK: %[[VAL_2545:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2546:.*]] = add nuw nsw i32 %[[VAL_2544]], %[[VAL_2545]] |
| // CHECK: %[[VAL_2547:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2548:.*]] = add nuw nsw i32 %[[VAL_2546]], %[[VAL_2547]] |
| // CHECK: %[[VAL_2549:.*]] = udiv i32 %[[VAL_2548]], 1 |
| // CHECK: %[[VAL_2550:.*]] = urem i32 %[[VAL_2549]], 32 |
| // CHECK: %[[VAL_2551:.*]] = udiv i32 %[[VAL_2548]], 32 |
| // CHECK: %[[VAL_2552:.*]] = urem i32 %[[VAL_2551]], 32 |
| // CHECK: %[[VAL_2553:.*]] = udiv i32 %[[VAL_2548]], 1024 |
| // CHECK: %[[VAL_2554:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2555:.*]] = getelementptr inbounds float, float* %[[VAL_2554]], i32 %[[VAL_2548]] |
| // CHECK: %[[VAL_2556:.*]] = load float, float* %[[VAL_2555]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2556]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2557:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2557]], float* %[[VAL_26]], float* %[[VAL_2557]]) |
| // CHECK: %[[VAL_2558:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2559:.*]] = getelementptr inbounds float, float* %[[VAL_2558]], i32 %[[VAL_2548]] |
| // CHECK: %[[VAL_2560:.*]] = load float, float* %[[VAL_2559]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2560]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2561:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2561]], float* %[[VAL_24]], float* %[[VAL_2561]]) |
| // CHECK: br label %[[VAL_1682]] |
| // CHECK: output_x_in_tile-true600: ; preds = %[[VAL_1682]] |
| // CHECK: %[[VAL_2562:.*]] = mul nuw nsw i32 %[[VAL_1684]], 1 |
| // CHECK: %[[VAL_2563:.*]] = add nuw nsw i32 0, %[[VAL_2562]] |
| // CHECK: %[[VAL_2564:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2565:.*]] = add nuw nsw i32 %[[VAL_2563]], %[[VAL_2564]] |
| // CHECK: %[[VAL_2566:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2567:.*]] = add nuw nsw i32 %[[VAL_2565]], %[[VAL_2566]] |
| // CHECK: %[[VAL_2568:.*]] = udiv i32 %[[VAL_2567]], 1 |
| // CHECK: %[[VAL_2569:.*]] = urem i32 %[[VAL_2568]], 32 |
| // CHECK: %[[VAL_2570:.*]] = udiv i32 %[[VAL_2567]], 32 |
| // CHECK: %[[VAL_2571:.*]] = urem i32 %[[VAL_2570]], 32 |
| // CHECK: %[[VAL_2572:.*]] = udiv i32 %[[VAL_2567]], 1024 |
| // CHECK: %[[VAL_2573:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2574:.*]] = getelementptr inbounds float, float* %[[VAL_2573]], i32 %[[VAL_2567]] |
| // CHECK: %[[VAL_2575:.*]] = load float, float* %[[VAL_2574]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2575]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2576:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2576]], float* %[[VAL_26]], float* %[[VAL_2576]]) |
| // CHECK: %[[VAL_2577:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2578:.*]] = getelementptr inbounds float, float* %[[VAL_2577]], i32 %[[VAL_2567]] |
| // CHECK: %[[VAL_2579:.*]] = load float, float* %[[VAL_2578]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2579]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2580:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2580]], float* %[[VAL_24]], float* %[[VAL_2580]]) |
| // CHECK: br label %[[VAL_1687]] |
| // CHECK: output_x_in_tile-true607: ; preds = %[[VAL_1687]] |
| // CHECK: %[[VAL_2581:.*]] = mul nuw nsw i32 %[[VAL_1689]], 1 |
| // CHECK: %[[VAL_2582:.*]] = add nuw nsw i32 0, %[[VAL_2581]] |
| // CHECK: %[[VAL_2583:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2584:.*]] = add nuw nsw i32 %[[VAL_2582]], %[[VAL_2583]] |
| // CHECK: %[[VAL_2585:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2586:.*]] = add nuw nsw i32 %[[VAL_2584]], %[[VAL_2585]] |
| // CHECK: %[[VAL_2587:.*]] = udiv i32 %[[VAL_2586]], 1 |
| // CHECK: %[[VAL_2588:.*]] = urem i32 %[[VAL_2587]], 32 |
| // CHECK: %[[VAL_2589:.*]] = udiv i32 %[[VAL_2586]], 32 |
| // CHECK: %[[VAL_2590:.*]] = urem i32 %[[VAL_2589]], 32 |
| // CHECK: %[[VAL_2591:.*]] = udiv i32 %[[VAL_2586]], 1024 |
| // CHECK: %[[VAL_2592:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2593:.*]] = getelementptr inbounds float, float* %[[VAL_2592]], i32 %[[VAL_2586]] |
| // CHECK: %[[VAL_2594:.*]] = load float, float* %[[VAL_2593]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2594]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2595:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2595]], float* %[[VAL_26]], float* %[[VAL_2595]]) |
| // CHECK: %[[VAL_2596:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2597:.*]] = getelementptr inbounds float, float* %[[VAL_2596]], i32 %[[VAL_2586]] |
| // CHECK: %[[VAL_2598:.*]] = load float, float* %[[VAL_2597]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2598]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2599:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2599]], float* %[[VAL_24]], float* %[[VAL_2599]]) |
| // CHECK: br label %[[VAL_1692]] |
| // CHECK: output_x_in_tile-true614: ; preds = %[[VAL_1692]] |
| // CHECK: %[[VAL_2600:.*]] = mul nuw nsw i32 %[[VAL_1694]], 1 |
| // CHECK: %[[VAL_2601:.*]] = add nuw nsw i32 0, %[[VAL_2600]] |
| // CHECK: %[[VAL_2602:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2603:.*]] = add nuw nsw i32 %[[VAL_2601]], %[[VAL_2602]] |
| // CHECK: %[[VAL_2604:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2605:.*]] = add nuw nsw i32 %[[VAL_2603]], %[[VAL_2604]] |
| // CHECK: %[[VAL_2606:.*]] = udiv i32 %[[VAL_2605]], 1 |
| // CHECK: %[[VAL_2607:.*]] = urem i32 %[[VAL_2606]], 32 |
| // CHECK: %[[VAL_2608:.*]] = udiv i32 %[[VAL_2605]], 32 |
| // CHECK: %[[VAL_2609:.*]] = urem i32 %[[VAL_2608]], 32 |
| // CHECK: %[[VAL_2610:.*]] = udiv i32 %[[VAL_2605]], 1024 |
| // CHECK: %[[VAL_2611:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2612:.*]] = getelementptr inbounds float, float* %[[VAL_2611]], i32 %[[VAL_2605]] |
| // CHECK: %[[VAL_2613:.*]] = load float, float* %[[VAL_2612]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2613]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2614:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2614]], float* %[[VAL_26]], float* %[[VAL_2614]]) |
| // CHECK: %[[VAL_2615:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2616:.*]] = getelementptr inbounds float, float* %[[VAL_2615]], i32 %[[VAL_2605]] |
| // CHECK: %[[VAL_2617:.*]] = load float, float* %[[VAL_2616]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2617]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2618:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2618]], float* %[[VAL_24]], float* %[[VAL_2618]]) |
| // CHECK: br label %[[VAL_1697]] |
| // CHECK: output_x_in_tile-true621: ; preds = %[[VAL_1697]] |
| // CHECK: %[[VAL_2619:.*]] = mul nuw nsw i32 %[[VAL_1699]], 1 |
| // CHECK: %[[VAL_2620:.*]] = add nuw nsw i32 0, %[[VAL_2619]] |
| // CHECK: %[[VAL_2621:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2622:.*]] = add nuw nsw i32 %[[VAL_2620]], %[[VAL_2621]] |
| // CHECK: %[[VAL_2623:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2624:.*]] = add nuw nsw i32 %[[VAL_2622]], %[[VAL_2623]] |
| // CHECK: %[[VAL_2625:.*]] = udiv i32 %[[VAL_2624]], 1 |
| // CHECK: %[[VAL_2626:.*]] = urem i32 %[[VAL_2625]], 32 |
| // CHECK: %[[VAL_2627:.*]] = udiv i32 %[[VAL_2624]], 32 |
| // CHECK: %[[VAL_2628:.*]] = urem i32 %[[VAL_2627]], 32 |
| // CHECK: %[[VAL_2629:.*]] = udiv i32 %[[VAL_2624]], 1024 |
| // CHECK: %[[VAL_2630:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2631:.*]] = getelementptr inbounds float, float* %[[VAL_2630]], i32 %[[VAL_2624]] |
| // CHECK: %[[VAL_2632:.*]] = load float, float* %[[VAL_2631]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2632]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2633:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2633]], float* %[[VAL_26]], float* %[[VAL_2633]]) |
| // CHECK: %[[VAL_2634:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2635:.*]] = getelementptr inbounds float, float* %[[VAL_2634]], i32 %[[VAL_2624]] |
| // CHECK: %[[VAL_2636:.*]] = load float, float* %[[VAL_2635]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2636]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2637:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2637]], float* %[[VAL_24]], float* %[[VAL_2637]]) |
| // CHECK: br label %[[VAL_1702]] |
| // CHECK: output_x_in_tile-true628: ; preds = %[[VAL_1702]] |
| // CHECK: %[[VAL_2638:.*]] = mul nuw nsw i32 %[[VAL_1704]], 1 |
| // CHECK: %[[VAL_2639:.*]] = add nuw nsw i32 0, %[[VAL_2638]] |
| // CHECK: %[[VAL_2640:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2641:.*]] = add nuw nsw i32 %[[VAL_2639]], %[[VAL_2640]] |
| // CHECK: %[[VAL_2642:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2643:.*]] = add nuw nsw i32 %[[VAL_2641]], %[[VAL_2642]] |
| // CHECK: %[[VAL_2644:.*]] = udiv i32 %[[VAL_2643]], 1 |
| // CHECK: %[[VAL_2645:.*]] = urem i32 %[[VAL_2644]], 32 |
| // CHECK: %[[VAL_2646:.*]] = udiv i32 %[[VAL_2643]], 32 |
| // CHECK: %[[VAL_2647:.*]] = urem i32 %[[VAL_2646]], 32 |
| // CHECK: %[[VAL_2648:.*]] = udiv i32 %[[VAL_2643]], 1024 |
| // CHECK: %[[VAL_2649:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2650:.*]] = getelementptr inbounds float, float* %[[VAL_2649]], i32 %[[VAL_2643]] |
| // CHECK: %[[VAL_2651:.*]] = load float, float* %[[VAL_2650]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2651]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2652:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2652]], float* %[[VAL_26]], float* %[[VAL_2652]]) |
| // CHECK: %[[VAL_2653:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2654:.*]] = getelementptr inbounds float, float* %[[VAL_2653]], i32 %[[VAL_2643]] |
| // CHECK: %[[VAL_2655:.*]] = load float, float* %[[VAL_2654]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2655]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2656:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2656]], float* %[[VAL_24]], float* %[[VAL_2656]]) |
| // CHECK: br label %[[VAL_1707]] |
| // CHECK: output_x_in_tile-true635: ; preds = %[[VAL_1707]] |
| // CHECK: %[[VAL_2657:.*]] = mul nuw nsw i32 %[[VAL_1709]], 1 |
| // CHECK: %[[VAL_2658:.*]] = add nuw nsw i32 0, %[[VAL_2657]] |
| // CHECK: %[[VAL_2659:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2660:.*]] = add nuw nsw i32 %[[VAL_2658]], %[[VAL_2659]] |
| // CHECK: %[[VAL_2661:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2662:.*]] = add nuw nsw i32 %[[VAL_2660]], %[[VAL_2661]] |
| // CHECK: %[[VAL_2663:.*]] = udiv i32 %[[VAL_2662]], 1 |
| // CHECK: %[[VAL_2664:.*]] = urem i32 %[[VAL_2663]], 32 |
| // CHECK: %[[VAL_2665:.*]] = udiv i32 %[[VAL_2662]], 32 |
| // CHECK: %[[VAL_2666:.*]] = urem i32 %[[VAL_2665]], 32 |
| // CHECK: %[[VAL_2667:.*]] = udiv i32 %[[VAL_2662]], 1024 |
| // CHECK: %[[VAL_2668:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2669:.*]] = getelementptr inbounds float, float* %[[VAL_2668]], i32 %[[VAL_2662]] |
| // CHECK: %[[VAL_2670:.*]] = load float, float* %[[VAL_2669]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2670]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2671:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2671]], float* %[[VAL_26]], float* %[[VAL_2671]]) |
| // CHECK: %[[VAL_2672:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2673:.*]] = getelementptr inbounds float, float* %[[VAL_2672]], i32 %[[VAL_2662]] |
| // CHECK: %[[VAL_2674:.*]] = load float, float* %[[VAL_2673]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2674]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2675:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2675]], float* %[[VAL_24]], float* %[[VAL_2675]]) |
| // CHECK: br label %[[VAL_1712]] |
| // CHECK: output_x_in_tile-true642: ; preds = %[[VAL_1712]] |
| // CHECK: %[[VAL_2676:.*]] = mul nuw nsw i32 %[[VAL_1714]], 1 |
| // CHECK: %[[VAL_2677:.*]] = add nuw nsw i32 0, %[[VAL_2676]] |
| // CHECK: %[[VAL_2678:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2679:.*]] = add nuw nsw i32 %[[VAL_2677]], %[[VAL_2678]] |
| // CHECK: %[[VAL_2680:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2681:.*]] = add nuw nsw i32 %[[VAL_2679]], %[[VAL_2680]] |
| // CHECK: %[[VAL_2682:.*]] = udiv i32 %[[VAL_2681]], 1 |
| // CHECK: %[[VAL_2683:.*]] = urem i32 %[[VAL_2682]], 32 |
| // CHECK: %[[VAL_2684:.*]] = udiv i32 %[[VAL_2681]], 32 |
| // CHECK: %[[VAL_2685:.*]] = urem i32 %[[VAL_2684]], 32 |
| // CHECK: %[[VAL_2686:.*]] = udiv i32 %[[VAL_2681]], 1024 |
| // CHECK: %[[VAL_2687:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2688:.*]] = getelementptr inbounds float, float* %[[VAL_2687]], i32 %[[VAL_2681]] |
| // CHECK: %[[VAL_2689:.*]] = load float, float* %[[VAL_2688]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2689]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2690:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2690]], float* %[[VAL_26]], float* %[[VAL_2690]]) |
| // CHECK: %[[VAL_2691:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2692:.*]] = getelementptr inbounds float, float* %[[VAL_2691]], i32 %[[VAL_2681]] |
| // CHECK: %[[VAL_2693:.*]] = load float, float* %[[VAL_2692]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2693]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2694:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2694]], float* %[[VAL_24]], float* %[[VAL_2694]]) |
| // CHECK: br label %[[VAL_1717]] |
| // CHECK: output_x_in_tile-true649: ; preds = %[[VAL_1717]] |
| // CHECK: %[[VAL_2695:.*]] = mul nuw nsw i32 %[[VAL_1719]], 1 |
| // CHECK: %[[VAL_2696:.*]] = add nuw nsw i32 0, %[[VAL_2695]] |
| // CHECK: %[[VAL_2697:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2698:.*]] = add nuw nsw i32 %[[VAL_2696]], %[[VAL_2697]] |
| // CHECK: %[[VAL_2699:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2700:.*]] = add nuw nsw i32 %[[VAL_2698]], %[[VAL_2699]] |
| // CHECK: %[[VAL_2701:.*]] = udiv i32 %[[VAL_2700]], 1 |
| // CHECK: %[[VAL_2702:.*]] = urem i32 %[[VAL_2701]], 32 |
| // CHECK: %[[VAL_2703:.*]] = udiv i32 %[[VAL_2700]], 32 |
| // CHECK: %[[VAL_2704:.*]] = urem i32 %[[VAL_2703]], 32 |
| // CHECK: %[[VAL_2705:.*]] = udiv i32 %[[VAL_2700]], 1024 |
| // CHECK: %[[VAL_2706:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2707:.*]] = getelementptr inbounds float, float* %[[VAL_2706]], i32 %[[VAL_2700]] |
| // CHECK: %[[VAL_2708:.*]] = load float, float* %[[VAL_2707]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2708]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2709:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2709]], float* %[[VAL_26]], float* %[[VAL_2709]]) |
| // CHECK: %[[VAL_2710:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2711:.*]] = getelementptr inbounds float, float* %[[VAL_2710]], i32 %[[VAL_2700]] |
| // CHECK: %[[VAL_2712:.*]] = load float, float* %[[VAL_2711]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2712]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2713:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2713]], float* %[[VAL_24]], float* %[[VAL_2713]]) |
| // CHECK: br label %[[VAL_1722]] |
| // CHECK: output_x_in_tile-true656: ; preds = %[[VAL_1722]] |
| // CHECK: %[[VAL_2714:.*]] = mul nuw nsw i32 %[[VAL_1724]], 1 |
| // CHECK: %[[VAL_2715:.*]] = add nuw nsw i32 0, %[[VAL_2714]] |
| // CHECK: %[[VAL_2716:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2717:.*]] = add nuw nsw i32 %[[VAL_2715]], %[[VAL_2716]] |
| // CHECK: %[[VAL_2718:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2719:.*]] = add nuw nsw i32 %[[VAL_2717]], %[[VAL_2718]] |
| // CHECK: %[[VAL_2720:.*]] = udiv i32 %[[VAL_2719]], 1 |
| // CHECK: %[[VAL_2721:.*]] = urem i32 %[[VAL_2720]], 32 |
| // CHECK: %[[VAL_2722:.*]] = udiv i32 %[[VAL_2719]], 32 |
| // CHECK: %[[VAL_2723:.*]] = urem i32 %[[VAL_2722]], 32 |
| // CHECK: %[[VAL_2724:.*]] = udiv i32 %[[VAL_2719]], 1024 |
| // CHECK: %[[VAL_2725:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2726:.*]] = getelementptr inbounds float, float* %[[VAL_2725]], i32 %[[VAL_2719]] |
| // CHECK: %[[VAL_2727:.*]] = load float, float* %[[VAL_2726]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2727]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2728:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2728]], float* %[[VAL_26]], float* %[[VAL_2728]]) |
| // CHECK: %[[VAL_2729:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2730:.*]] = getelementptr inbounds float, float* %[[VAL_2729]], i32 %[[VAL_2719]] |
| // CHECK: %[[VAL_2731:.*]] = load float, float* %[[VAL_2730]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2731]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2732:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2732]], float* %[[VAL_24]], float* %[[VAL_2732]]) |
| // CHECK: br label %[[VAL_1727]] |
| // CHECK: output_x_in_tile-true663: ; preds = %[[VAL_1727]] |
| // CHECK: %[[VAL_2733:.*]] = mul nuw nsw i32 %[[VAL_1729]], 1 |
| // CHECK: %[[VAL_2734:.*]] = add nuw nsw i32 0, %[[VAL_2733]] |
| // CHECK: %[[VAL_2735:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2736:.*]] = add nuw nsw i32 %[[VAL_2734]], %[[VAL_2735]] |
| // CHECK: %[[VAL_2737:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2738:.*]] = add nuw nsw i32 %[[VAL_2736]], %[[VAL_2737]] |
| // CHECK: %[[VAL_2739:.*]] = udiv i32 %[[VAL_2738]], 1 |
| // CHECK: %[[VAL_2740:.*]] = urem i32 %[[VAL_2739]], 32 |
| // CHECK: %[[VAL_2741:.*]] = udiv i32 %[[VAL_2738]], 32 |
| // CHECK: %[[VAL_2742:.*]] = urem i32 %[[VAL_2741]], 32 |
| // CHECK: %[[VAL_2743:.*]] = udiv i32 %[[VAL_2738]], 1024 |
| // CHECK: %[[VAL_2744:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2745:.*]] = getelementptr inbounds float, float* %[[VAL_2744]], i32 %[[VAL_2738]] |
| // CHECK: %[[VAL_2746:.*]] = load float, float* %[[VAL_2745]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2746]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2747:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2747]], float* %[[VAL_26]], float* %[[VAL_2747]]) |
| // CHECK: %[[VAL_2748:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2749:.*]] = getelementptr inbounds float, float* %[[VAL_2748]], i32 %[[VAL_2738]] |
| // CHECK: %[[VAL_2750:.*]] = load float, float* %[[VAL_2749]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2750]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2751:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2751]], float* %[[VAL_24]], float* %[[VAL_2751]]) |
| // CHECK: br label %[[VAL_1732]] |
| // CHECK: output_x_in_tile-true670: ; preds = %[[VAL_1732]] |
| // CHECK: %[[VAL_2752:.*]] = mul nuw nsw i32 %[[VAL_1734]], 1 |
| // CHECK: %[[VAL_2753:.*]] = add nuw nsw i32 0, %[[VAL_2752]] |
| // CHECK: %[[VAL_2754:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2755:.*]] = add nuw nsw i32 %[[VAL_2753]], %[[VAL_2754]] |
| // CHECK: %[[VAL_2756:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2757:.*]] = add nuw nsw i32 %[[VAL_2755]], %[[VAL_2756]] |
| // CHECK: %[[VAL_2758:.*]] = udiv i32 %[[VAL_2757]], 1 |
| // CHECK: %[[VAL_2759:.*]] = urem i32 %[[VAL_2758]], 32 |
| // CHECK: %[[VAL_2760:.*]] = udiv i32 %[[VAL_2757]], 32 |
| // CHECK: %[[VAL_2761:.*]] = urem i32 %[[VAL_2760]], 32 |
| // CHECK: %[[VAL_2762:.*]] = udiv i32 %[[VAL_2757]], 1024 |
| // CHECK: %[[VAL_2763:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2764:.*]] = getelementptr inbounds float, float* %[[VAL_2763]], i32 %[[VAL_2757]] |
| // CHECK: %[[VAL_2765:.*]] = load float, float* %[[VAL_2764]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2765]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2766:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2766]], float* %[[VAL_26]], float* %[[VAL_2766]]) |
| // CHECK: %[[VAL_2767:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2768:.*]] = getelementptr inbounds float, float* %[[VAL_2767]], i32 %[[VAL_2757]] |
| // CHECK: %[[VAL_2769:.*]] = load float, float* %[[VAL_2768]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2769]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2770:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2770]], float* %[[VAL_24]], float* %[[VAL_2770]]) |
| // CHECK: br label %[[VAL_1737]] |
| // CHECK: output_x_in_tile-true677: ; preds = %[[VAL_1737]] |
| // CHECK: %[[VAL_2771:.*]] = mul nuw nsw i32 %[[VAL_1739]], 1 |
| // CHECK: %[[VAL_2772:.*]] = add nuw nsw i32 0, %[[VAL_2771]] |
| // CHECK: %[[VAL_2773:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2774:.*]] = add nuw nsw i32 %[[VAL_2772]], %[[VAL_2773]] |
| // CHECK: %[[VAL_2775:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2776:.*]] = add nuw nsw i32 %[[VAL_2774]], %[[VAL_2775]] |
| // CHECK: %[[VAL_2777:.*]] = udiv i32 %[[VAL_2776]], 1 |
| // CHECK: %[[VAL_2778:.*]] = urem i32 %[[VAL_2777]], 32 |
| // CHECK: %[[VAL_2779:.*]] = udiv i32 %[[VAL_2776]], 32 |
| // CHECK: %[[VAL_2780:.*]] = urem i32 %[[VAL_2779]], 32 |
| // CHECK: %[[VAL_2781:.*]] = udiv i32 %[[VAL_2776]], 1024 |
| // CHECK: %[[VAL_2782:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2783:.*]] = getelementptr inbounds float, float* %[[VAL_2782]], i32 %[[VAL_2776]] |
| // CHECK: %[[VAL_2784:.*]] = load float, float* %[[VAL_2783]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2784]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2785:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2785]], float* %[[VAL_26]], float* %[[VAL_2785]]) |
| // CHECK: %[[VAL_2786:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2787:.*]] = getelementptr inbounds float, float* %[[VAL_2786]], i32 %[[VAL_2776]] |
| // CHECK: %[[VAL_2788:.*]] = load float, float* %[[VAL_2787]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2788]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2789:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2789]], float* %[[VAL_24]], float* %[[VAL_2789]]) |
| // CHECK: br label %[[VAL_1742]] |
| // CHECK: output_x_in_tile-true684: ; preds = %[[VAL_1742]] |
| // CHECK: %[[VAL_2790:.*]] = mul nuw nsw i32 %[[VAL_1744]], 1 |
| // CHECK: %[[VAL_2791:.*]] = add nuw nsw i32 0, %[[VAL_2790]] |
| // CHECK: %[[VAL_2792:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2793:.*]] = add nuw nsw i32 %[[VAL_2791]], %[[VAL_2792]] |
| // CHECK: %[[VAL_2794:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2795:.*]] = add nuw nsw i32 %[[VAL_2793]], %[[VAL_2794]] |
| // CHECK: %[[VAL_2796:.*]] = udiv i32 %[[VAL_2795]], 1 |
| // CHECK: %[[VAL_2797:.*]] = urem i32 %[[VAL_2796]], 32 |
| // CHECK: %[[VAL_2798:.*]] = udiv i32 %[[VAL_2795]], 32 |
| // CHECK: %[[VAL_2799:.*]] = urem i32 %[[VAL_2798]], 32 |
| // CHECK: %[[VAL_2800:.*]] = udiv i32 %[[VAL_2795]], 1024 |
| // CHECK: %[[VAL_2801:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2802:.*]] = getelementptr inbounds float, float* %[[VAL_2801]], i32 %[[VAL_2795]] |
| // CHECK: %[[VAL_2803:.*]] = load float, float* %[[VAL_2802]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2803]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2804:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2804]], float* %[[VAL_26]], float* %[[VAL_2804]]) |
| // CHECK: %[[VAL_2805:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2806:.*]] = getelementptr inbounds float, float* %[[VAL_2805]], i32 %[[VAL_2795]] |
| // CHECK: %[[VAL_2807:.*]] = load float, float* %[[VAL_2806]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2807]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2808:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2808]], float* %[[VAL_24]], float* %[[VAL_2808]]) |
| // CHECK: br label %[[VAL_1747]] |
| // CHECK: output_x_in_tile-true691: ; preds = %[[VAL_1747]] |
| // CHECK: %[[VAL_2809:.*]] = mul nuw nsw i32 %[[VAL_1749]], 1 |
| // CHECK: %[[VAL_2810:.*]] = add nuw nsw i32 0, %[[VAL_2809]] |
| // CHECK: %[[VAL_2811:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2812:.*]] = add nuw nsw i32 %[[VAL_2810]], %[[VAL_2811]] |
| // CHECK: %[[VAL_2813:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2814:.*]] = add nuw nsw i32 %[[VAL_2812]], %[[VAL_2813]] |
| // CHECK: %[[VAL_2815:.*]] = udiv i32 %[[VAL_2814]], 1 |
| // CHECK: %[[VAL_2816:.*]] = urem i32 %[[VAL_2815]], 32 |
| // CHECK: %[[VAL_2817:.*]] = udiv i32 %[[VAL_2814]], 32 |
| // CHECK: %[[VAL_2818:.*]] = urem i32 %[[VAL_2817]], 32 |
| // CHECK: %[[VAL_2819:.*]] = udiv i32 %[[VAL_2814]], 1024 |
| // CHECK: %[[VAL_2820:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2821:.*]] = getelementptr inbounds float, float* %[[VAL_2820]], i32 %[[VAL_2814]] |
| // CHECK: %[[VAL_2822:.*]] = load float, float* %[[VAL_2821]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2822]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2823:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2823]], float* %[[VAL_26]], float* %[[VAL_2823]]) |
| // CHECK: %[[VAL_2824:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2825:.*]] = getelementptr inbounds float, float* %[[VAL_2824]], i32 %[[VAL_2814]] |
| // CHECK: %[[VAL_2826:.*]] = load float, float* %[[VAL_2825]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2826]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2827:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2827]], float* %[[VAL_24]], float* %[[VAL_2827]]) |
| // CHECK: br label %[[VAL_1752]] |
| // CHECK: output_x_in_tile-true698: ; preds = %[[VAL_1752]] |
| // CHECK: %[[VAL_2828:.*]] = mul nuw nsw i32 %[[VAL_1754]], 1 |
| // CHECK: %[[VAL_2829:.*]] = add nuw nsw i32 0, %[[VAL_2828]] |
| // CHECK: %[[VAL_2830:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2831:.*]] = add nuw nsw i32 %[[VAL_2829]], %[[VAL_2830]] |
| // CHECK: %[[VAL_2832:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2833:.*]] = add nuw nsw i32 %[[VAL_2831]], %[[VAL_2832]] |
| // CHECK: %[[VAL_2834:.*]] = udiv i32 %[[VAL_2833]], 1 |
| // CHECK: %[[VAL_2835:.*]] = urem i32 %[[VAL_2834]], 32 |
| // CHECK: %[[VAL_2836:.*]] = udiv i32 %[[VAL_2833]], 32 |
| // CHECK: %[[VAL_2837:.*]] = urem i32 %[[VAL_2836]], 32 |
| // CHECK: %[[VAL_2838:.*]] = udiv i32 %[[VAL_2833]], 1024 |
| // CHECK: %[[VAL_2839:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2840:.*]] = getelementptr inbounds float, float* %[[VAL_2839]], i32 %[[VAL_2833]] |
| // CHECK: %[[VAL_2841:.*]] = load float, float* %[[VAL_2840]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2841]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2842:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2842]], float* %[[VAL_26]], float* %[[VAL_2842]]) |
| // CHECK: %[[VAL_2843:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2844:.*]] = getelementptr inbounds float, float* %[[VAL_2843]], i32 %[[VAL_2833]] |
| // CHECK: %[[VAL_2845:.*]] = load float, float* %[[VAL_2844]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2845]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2846:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2846]], float* %[[VAL_24]], float* %[[VAL_2846]]) |
| // CHECK: br label %[[VAL_1757]] |
| // CHECK: output_x_in_tile-true705: ; preds = %[[VAL_1757]] |
| // CHECK: %[[VAL_2847:.*]] = mul nuw nsw i32 %[[VAL_1759]], 1 |
| // CHECK: %[[VAL_2848:.*]] = add nuw nsw i32 0, %[[VAL_2847]] |
| // CHECK: %[[VAL_2849:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2850:.*]] = add nuw nsw i32 %[[VAL_2848]], %[[VAL_2849]] |
| // CHECK: %[[VAL_2851:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2852:.*]] = add nuw nsw i32 %[[VAL_2850]], %[[VAL_2851]] |
| // CHECK: %[[VAL_2853:.*]] = udiv i32 %[[VAL_2852]], 1 |
| // CHECK: %[[VAL_2854:.*]] = urem i32 %[[VAL_2853]], 32 |
| // CHECK: %[[VAL_2855:.*]] = udiv i32 %[[VAL_2852]], 32 |
| // CHECK: %[[VAL_2856:.*]] = urem i32 %[[VAL_2855]], 32 |
| // CHECK: %[[VAL_2857:.*]] = udiv i32 %[[VAL_2852]], 1024 |
| // CHECK: %[[VAL_2858:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2859:.*]] = getelementptr inbounds float, float* %[[VAL_2858]], i32 %[[VAL_2852]] |
| // CHECK: %[[VAL_2860:.*]] = load float, float* %[[VAL_2859]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2860]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2861:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2861]], float* %[[VAL_26]], float* %[[VAL_2861]]) |
| // CHECK: %[[VAL_2862:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2863:.*]] = getelementptr inbounds float, float* %[[VAL_2862]], i32 %[[VAL_2852]] |
| // CHECK: %[[VAL_2864:.*]] = load float, float* %[[VAL_2863]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2864]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2865:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2865]], float* %[[VAL_24]], float* %[[VAL_2865]]) |
| // CHECK: br label %[[VAL_1762]] |
| // CHECK: output_x_in_tile-true712: ; preds = %[[VAL_1762]] |
| // CHECK: %[[VAL_2866:.*]] = mul nuw nsw i32 %[[VAL_1764]], 1 |
| // CHECK: %[[VAL_2867:.*]] = add nuw nsw i32 0, %[[VAL_2866]] |
| // CHECK: %[[VAL_2868:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2869:.*]] = add nuw nsw i32 %[[VAL_2867]], %[[VAL_2868]] |
| // CHECK: %[[VAL_2870:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2871:.*]] = add nuw nsw i32 %[[VAL_2869]], %[[VAL_2870]] |
| // CHECK: %[[VAL_2872:.*]] = udiv i32 %[[VAL_2871]], 1 |
| // CHECK: %[[VAL_2873:.*]] = urem i32 %[[VAL_2872]], 32 |
| // CHECK: %[[VAL_2874:.*]] = udiv i32 %[[VAL_2871]], 32 |
| // CHECK: %[[VAL_2875:.*]] = urem i32 %[[VAL_2874]], 32 |
| // CHECK: %[[VAL_2876:.*]] = udiv i32 %[[VAL_2871]], 1024 |
| // CHECK: %[[VAL_2877:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2878:.*]] = getelementptr inbounds float, float* %[[VAL_2877]], i32 %[[VAL_2871]] |
| // CHECK: %[[VAL_2879:.*]] = load float, float* %[[VAL_2878]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2879]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2880:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2880]], float* %[[VAL_26]], float* %[[VAL_2880]]) |
| // CHECK: %[[VAL_2881:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2882:.*]] = getelementptr inbounds float, float* %[[VAL_2881]], i32 %[[VAL_2871]] |
| // CHECK: %[[VAL_2883:.*]] = load float, float* %[[VAL_2882]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2883]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2884:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2884]], float* %[[VAL_24]], float* %[[VAL_2884]]) |
| // CHECK: br label %[[VAL_1767]] |
| // CHECK: output_x_in_tile-true719: ; preds = %[[VAL_1767]] |
| // CHECK: %[[VAL_2885:.*]] = mul nuw nsw i32 %[[VAL_1769]], 1 |
| // CHECK: %[[VAL_2886:.*]] = add nuw nsw i32 0, %[[VAL_2885]] |
| // CHECK: %[[VAL_2887:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2888:.*]] = add nuw nsw i32 %[[VAL_2886]], %[[VAL_2887]] |
| // CHECK: %[[VAL_2889:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2890:.*]] = add nuw nsw i32 %[[VAL_2888]], %[[VAL_2889]] |
| // CHECK: %[[VAL_2891:.*]] = udiv i32 %[[VAL_2890]], 1 |
| // CHECK: %[[VAL_2892:.*]] = urem i32 %[[VAL_2891]], 32 |
| // CHECK: %[[VAL_2893:.*]] = udiv i32 %[[VAL_2890]], 32 |
| // CHECK: %[[VAL_2894:.*]] = urem i32 %[[VAL_2893]], 32 |
| // CHECK: %[[VAL_2895:.*]] = udiv i32 %[[VAL_2890]], 1024 |
| // CHECK: %[[VAL_2896:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2897:.*]] = getelementptr inbounds float, float* %[[VAL_2896]], i32 %[[VAL_2890]] |
| // CHECK: %[[VAL_2898:.*]] = load float, float* %[[VAL_2897]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2898]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2899:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2899]], float* %[[VAL_26]], float* %[[VAL_2899]]) |
| // CHECK: %[[VAL_2900:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2901:.*]] = getelementptr inbounds float, float* %[[VAL_2900]], i32 %[[VAL_2890]] |
| // CHECK: %[[VAL_2902:.*]] = load float, float* %[[VAL_2901]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2902]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2903:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2903]], float* %[[VAL_24]], float* %[[VAL_2903]]) |
| // CHECK: br label %[[VAL_1772]] |
| // CHECK: output_x_in_tile-true726: ; preds = %[[VAL_1772]] |
| // CHECK: %[[VAL_2904:.*]] = mul nuw nsw i32 %[[VAL_1774]], 1 |
| // CHECK: %[[VAL_2905:.*]] = add nuw nsw i32 0, %[[VAL_2904]] |
| // CHECK: %[[VAL_2906:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2907:.*]] = add nuw nsw i32 %[[VAL_2905]], %[[VAL_2906]] |
| // CHECK: %[[VAL_2908:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2909:.*]] = add nuw nsw i32 %[[VAL_2907]], %[[VAL_2908]] |
| // CHECK: %[[VAL_2910:.*]] = udiv i32 %[[VAL_2909]], 1 |
| // CHECK: %[[VAL_2911:.*]] = urem i32 %[[VAL_2910]], 32 |
| // CHECK: %[[VAL_2912:.*]] = udiv i32 %[[VAL_2909]], 32 |
| // CHECK: %[[VAL_2913:.*]] = urem i32 %[[VAL_2912]], 32 |
| // CHECK: %[[VAL_2914:.*]] = udiv i32 %[[VAL_2909]], 1024 |
| // CHECK: %[[VAL_2915:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2916:.*]] = getelementptr inbounds float, float* %[[VAL_2915]], i32 %[[VAL_2909]] |
| // CHECK: %[[VAL_2917:.*]] = load float, float* %[[VAL_2916]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2917]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2918:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2918]], float* %[[VAL_26]], float* %[[VAL_2918]]) |
| // CHECK: %[[VAL_2919:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2920:.*]] = getelementptr inbounds float, float* %[[VAL_2919]], i32 %[[VAL_2909]] |
| // CHECK: %[[VAL_2921:.*]] = load float, float* %[[VAL_2920]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2921]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2922:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2922]], float* %[[VAL_24]], float* %[[VAL_2922]]) |
| // CHECK: br label %[[VAL_1777]] |
| // CHECK: output_x_in_tile-true733: ; preds = %[[VAL_1777]] |
| // CHECK: %[[VAL_2923:.*]] = mul nuw nsw i32 %[[VAL_1779]], 1 |
| // CHECK: %[[VAL_2924:.*]] = add nuw nsw i32 0, %[[VAL_2923]] |
| // CHECK: %[[VAL_2925:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2926:.*]] = add nuw nsw i32 %[[VAL_2924]], %[[VAL_2925]] |
| // CHECK: %[[VAL_2927:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2928:.*]] = add nuw nsw i32 %[[VAL_2926]], %[[VAL_2927]] |
| // CHECK: %[[VAL_2929:.*]] = udiv i32 %[[VAL_2928]], 1 |
| // CHECK: %[[VAL_2930:.*]] = urem i32 %[[VAL_2929]], 32 |
| // CHECK: %[[VAL_2931:.*]] = udiv i32 %[[VAL_2928]], 32 |
| // CHECK: %[[VAL_2932:.*]] = urem i32 %[[VAL_2931]], 32 |
| // CHECK: %[[VAL_2933:.*]] = udiv i32 %[[VAL_2928]], 1024 |
| // CHECK: %[[VAL_2934:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2935:.*]] = getelementptr inbounds float, float* %[[VAL_2934]], i32 %[[VAL_2928]] |
| // CHECK: %[[VAL_2936:.*]] = load float, float* %[[VAL_2935]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2936]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2937:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2937]], float* %[[VAL_26]], float* %[[VAL_2937]]) |
| // CHECK: %[[VAL_2938:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2939:.*]] = getelementptr inbounds float, float* %[[VAL_2938]], i32 %[[VAL_2928]] |
| // CHECK: %[[VAL_2940:.*]] = load float, float* %[[VAL_2939]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2940]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2941:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2941]], float* %[[VAL_24]], float* %[[VAL_2941]]) |
| // CHECK: br label %[[VAL_1782]] |
| // CHECK: output_x_in_tile-true740: ; preds = %[[VAL_1782]] |
| // CHECK: %[[VAL_2942:.*]] = mul nuw nsw i32 %[[VAL_1784]], 1 |
| // CHECK: %[[VAL_2943:.*]] = add nuw nsw i32 0, %[[VAL_2942]] |
| // CHECK: %[[VAL_2944:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2945:.*]] = add nuw nsw i32 %[[VAL_2943]], %[[VAL_2944]] |
| // CHECK: %[[VAL_2946:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2947:.*]] = add nuw nsw i32 %[[VAL_2945]], %[[VAL_2946]] |
| // CHECK: %[[VAL_2948:.*]] = udiv i32 %[[VAL_2947]], 1 |
| // CHECK: %[[VAL_2949:.*]] = urem i32 %[[VAL_2948]], 32 |
| // CHECK: %[[VAL_2950:.*]] = udiv i32 %[[VAL_2947]], 32 |
| // CHECK: %[[VAL_2951:.*]] = urem i32 %[[VAL_2950]], 32 |
| // CHECK: %[[VAL_2952:.*]] = udiv i32 %[[VAL_2947]], 1024 |
| // CHECK: %[[VAL_2953:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2954:.*]] = getelementptr inbounds float, float* %[[VAL_2953]], i32 %[[VAL_2947]] |
| // CHECK: %[[VAL_2955:.*]] = load float, float* %[[VAL_2954]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2955]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2956:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2956]], float* %[[VAL_26]], float* %[[VAL_2956]]) |
| // CHECK: %[[VAL_2957:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2958:.*]] = getelementptr inbounds float, float* %[[VAL_2957]], i32 %[[VAL_2947]] |
| // CHECK: %[[VAL_2959:.*]] = load float, float* %[[VAL_2958]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2959]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2960:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2960]], float* %[[VAL_24]], float* %[[VAL_2960]]) |
| // CHECK: br label %[[VAL_1787]] |
| // CHECK: output_x_in_tile-true747: ; preds = %[[VAL_1787]] |
| // CHECK: %[[VAL_2961:.*]] = mul nuw nsw i32 %[[VAL_1789]], 1 |
| // CHECK: %[[VAL_2962:.*]] = add nuw nsw i32 0, %[[VAL_2961]] |
| // CHECK: %[[VAL_2963:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2964:.*]] = add nuw nsw i32 %[[VAL_2962]], %[[VAL_2963]] |
| // CHECK: %[[VAL_2965:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2966:.*]] = add nuw nsw i32 %[[VAL_2964]], %[[VAL_2965]] |
| // CHECK: %[[VAL_2967:.*]] = udiv i32 %[[VAL_2966]], 1 |
| // CHECK: %[[VAL_2968:.*]] = urem i32 %[[VAL_2967]], 32 |
| // CHECK: %[[VAL_2969:.*]] = udiv i32 %[[VAL_2966]], 32 |
| // CHECK: %[[VAL_2970:.*]] = urem i32 %[[VAL_2969]], 32 |
| // CHECK: %[[VAL_2971:.*]] = udiv i32 %[[VAL_2966]], 1024 |
| // CHECK: %[[VAL_2972:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2973:.*]] = getelementptr inbounds float, float* %[[VAL_2972]], i32 %[[VAL_2966]] |
| // CHECK: %[[VAL_2974:.*]] = load float, float* %[[VAL_2973]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2974]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2975:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2975]], float* %[[VAL_26]], float* %[[VAL_2975]]) |
| // CHECK: %[[VAL_2976:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2977:.*]] = getelementptr inbounds float, float* %[[VAL_2976]], i32 %[[VAL_2966]] |
| // CHECK: %[[VAL_2978:.*]] = load float, float* %[[VAL_2977]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2978]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2979:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2979]], float* %[[VAL_24]], float* %[[VAL_2979]]) |
| // CHECK: br label %[[VAL_1792]] |
| // CHECK: output_x_in_tile-true754: ; preds = %[[VAL_1792]] |
| // CHECK: %[[VAL_2980:.*]] = mul nuw nsw i32 %[[VAL_1794]], 1 |
| // CHECK: %[[VAL_2981:.*]] = add nuw nsw i32 0, %[[VAL_2980]] |
| // CHECK: %[[VAL_2982:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_2983:.*]] = add nuw nsw i32 %[[VAL_2981]], %[[VAL_2982]] |
| // CHECK: %[[VAL_2984:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_2985:.*]] = add nuw nsw i32 %[[VAL_2983]], %[[VAL_2984]] |
| // CHECK: %[[VAL_2986:.*]] = udiv i32 %[[VAL_2985]], 1 |
| // CHECK: %[[VAL_2987:.*]] = urem i32 %[[VAL_2986]], 32 |
| // CHECK: %[[VAL_2988:.*]] = udiv i32 %[[VAL_2985]], 32 |
| // CHECK: %[[VAL_2989:.*]] = urem i32 %[[VAL_2988]], 32 |
| // CHECK: %[[VAL_2990:.*]] = udiv i32 %[[VAL_2985]], 1024 |
| // CHECK: %[[VAL_2991:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2992:.*]] = getelementptr inbounds float, float* %[[VAL_2991]], i32 %[[VAL_2985]] |
| // CHECK: %[[VAL_2993:.*]] = load float, float* %[[VAL_2992]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2993]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_2994:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_2994]], float* %[[VAL_26]], float* %[[VAL_2994]]) |
| // CHECK: %[[VAL_2995:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_2996:.*]] = getelementptr inbounds float, float* %[[VAL_2995]], i32 %[[VAL_2985]] |
| // CHECK: %[[VAL_2997:.*]] = load float, float* %[[VAL_2996]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_2997]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_2998:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_2998]], float* %[[VAL_24]], float* %[[VAL_2998]]) |
| // CHECK: br label %[[VAL_1797]] |
| // CHECK: output_x_in_tile-true761: ; preds = %[[VAL_1797]] |
| // CHECK: %[[VAL_2999:.*]] = mul nuw nsw i32 %[[VAL_1799]], 1 |
| // CHECK: %[[VAL_3000:.*]] = add nuw nsw i32 0, %[[VAL_2999]] |
| // CHECK: %[[VAL_3001:.*]] = mul nuw nsw i32 %[[VAL_1482]], 32 |
| // CHECK: %[[VAL_3002:.*]] = add nuw nsw i32 %[[VAL_3000]], %[[VAL_3001]] |
| // CHECK: %[[VAL_3003:.*]] = mul nuw nsw i32 %[[VAL_62]], 2048 |
| // CHECK: %[[VAL_3004:.*]] = add nuw nsw i32 %[[VAL_3002]], %[[VAL_3003]] |
| // CHECK: %[[VAL_3005:.*]] = udiv i32 %[[VAL_3004]], 1 |
| // CHECK: %[[VAL_3006:.*]] = urem i32 %[[VAL_3005]], 32 |
| // CHECK: %[[VAL_3007:.*]] = udiv i32 %[[VAL_3004]], 32 |
| // CHECK: %[[VAL_3008:.*]] = urem i32 %[[VAL_3007]], 32 |
| // CHECK: %[[VAL_3009:.*]] = udiv i32 %[[VAL_3004]], 1024 |
| // CHECK: %[[VAL_3010:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_3011:.*]] = getelementptr inbounds float, float* %[[VAL_3010]], i32 %[[VAL_3004]] |
| // CHECK: %[[VAL_3012:.*]] = load float, float* %[[VAL_3011]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_3012]], float* %[[VAL_26]], align 4 |
| // CHECK: %[[VAL_3013:.*]] = getelementptr inbounds float, float* %[[VAL_25]], i32 0 |
| // CHECK: call void @region_1_4(float* %[[VAL_3013]], float* %[[VAL_26]], float* %[[VAL_3013]]) |
| // CHECK: %[[VAL_3014:.*]] = bitcast [2 x [32 x [32 x float]]]* %[[VAL_29]] to float* |
| // CHECK: %[[VAL_3015:.*]] = getelementptr inbounds float, float* %[[VAL_3014]], i32 %[[VAL_3004]] |
| // CHECK: %[[VAL_3016:.*]] = load float, float* %[[VAL_3015]], align 4, !invariant.load !3 |
| // CHECK: store float %[[VAL_3016]], float* %[[VAL_24]], align 4 |
| // CHECK: %[[VAL_3017:.*]] = getelementptr inbounds float, float* %[[VAL_23]], i32 0 |
| // CHECK: call void @region_2_9(float* %[[VAL_3017]], float* %[[VAL_24]], float* %[[VAL_3017]]) |
| // CHECK: br label %[[VAL_82]] |
| // CHECK: intra_warp_reduce_write-true: ; preds = %[[VAL_75]] |
| // CHECK: %[[VAL_3018:.*]] = getelementptr inbounds [1 x [32 x float]], [1 x [32 x float]] addrspace(3)* @shared_cache_0, i32 0, i32 0, i32 %[[VAL_106]] |
| // CHECK: %[[VAL_3019:.*]] = addrspacecast float addrspace(3)* %[[VAL_3018]] to float* |
| // CHECK: %[[VAL_3020:.*]] = load float, float* %[[VAL_95]], align 4 |
| // CHECK: store float %[[VAL_3020]], float* %[[VAL_3019]], align 4 |
| // CHECK: br label %[[VAL_109]] |
| // CHECK: inter_warp_reduce-true: ; preds = %[[VAL_109]] |
| // CHECK: %[[VAL_3021:.*]] = getelementptr inbounds [1 x [32 x float]], [1 x [32 x float]] addrspace(3)* @shared_cache_0, i32 0, i32 0, i32 %[[VAL_86]] |
| // CHECK: %[[VAL_3022:.*]] = addrspacecast float addrspace(3)* %[[VAL_3021]] to float* |
| // CHECK: store float %[[VAL_48]], float* %[[VAL_16]], align 4 |
| // CHECK: %[[VAL_3023:.*]] = icmp ult i32 %[[VAL_84]], 1 |
| // CHECK: %[[VAL_3024:.*]] = select i1 %[[VAL_3023]], float* %[[VAL_3022]], float* %[[VAL_16]] |
| // CHECK: %[[VAL_3025:.*]] = load float, float* %[[VAL_3024]], align 4 |
| // CHECK: %[[VAL_3026:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3025]], i32 16, i32 31) |
| // CHECK: store float %[[VAL_3026]], float* %[[VAL_15]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_3024]], float* %[[VAL_15]], float* %[[VAL_3024]]) |
| // CHECK: %[[VAL_3027:.*]] = load float, float* %[[VAL_3024]], align 4 |
| // CHECK: %[[VAL_3028:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3027]], i32 8, i32 31) |
| // CHECK: store float %[[VAL_3028]], float* %[[VAL_14]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_3024]], float* %[[VAL_14]], float* %[[VAL_3024]]) |
| // CHECK: %[[VAL_3029:.*]] = load float, float* %[[VAL_3024]], align 4 |
| // CHECK: %[[VAL_3030:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3029]], i32 4, i32 31) |
| // CHECK: store float %[[VAL_3030]], float* %[[VAL_13]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_3024]], float* %[[VAL_13]], float* %[[VAL_3024]]) |
| // CHECK: %[[VAL_3031:.*]] = load float, float* %[[VAL_3024]], align 4 |
| // CHECK: %[[VAL_3032:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3031]], i32 2, i32 31) |
| // CHECK: store float %[[VAL_3032]], float* %[[VAL_12]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_3024]], float* %[[VAL_12]], float* %[[VAL_3024]]) |
| // CHECK: %[[VAL_3033:.*]] = load float, float* %[[VAL_3024]], align 4 |
| // CHECK: %[[VAL_3034:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3033]], i32 1, i32 31) |
| // CHECK: store float %[[VAL_3034]], float* %[[VAL_11]], align 4 |
| // CHECK: call void @region_1_4(float* %[[VAL_3024]], float* %[[VAL_11]], float* %[[VAL_3024]]) |
| // CHECK: %[[VAL_3035:.*]] = icmp eq i32 %[[VAL_84]], 0 |
| // CHECK: br i1 %[[VAL_3035]], label %[[VAL_3036:.*]], label %[[VAL_113]] |
| // CHECK: reduction_write_output-after: ; preds = %[[VAL_3036]], %[[VAL_111]] |
| // CHECK: br label %[[VAL_112]] |
| // CHECK: reduction_write_output-true: ; preds = %[[VAL_111]] |
| // CHECK: %[[VAL_3037:.*]] = load float, float* %[[VAL_3022]], align 4 |
| // CHECK: store float %[[VAL_3037]], float* %[[VAL_94]], align 4 |
| // CHECK: br label %[[VAL_113]] |
| // CHECK: intra_warp_reduce_write-true800: ; preds = %[[VAL_112]] |
| // CHECK: %[[VAL_3038:.*]] = getelementptr inbounds [1 x [32 x float]], [1 x [32 x float]] addrspace(3)* @shared_cache_1, i32 0, i32 0, i32 %[[VAL_130]] |
| // CHECK: %[[VAL_3039:.*]] = addrspacecast float addrspace(3)* %[[VAL_3038]] to float* |
| // CHECK: %[[VAL_3040:.*]] = load float, float* %[[VAL_119]], align 4 |
| // CHECK: store float %[[VAL_3040]], float* %[[VAL_3039]], align 4 |
| // CHECK: br label %[[VAL_133]] |
| // CHECK: inter_warp_reduce-true802: ; preds = %[[VAL_133]] |
| // CHECK: %[[VAL_3041:.*]] = getelementptr inbounds [1 x [32 x float]], [1 x [32 x float]] addrspace(3)* @shared_cache_1, i32 0, i32 0, i32 %[[VAL_86]] |
| // CHECK: %[[VAL_3042:.*]] = addrspacecast float addrspace(3)* %[[VAL_3041]] to float* |
| // CHECK: store float %[[VAL_50]], float* %[[VAL_5]], align 4 |
| // CHECK: %[[VAL_3043:.*]] = icmp ult i32 %[[VAL_84]], 1 |
| // CHECK: %[[VAL_3044:.*]] = select i1 %[[VAL_3043]], float* %[[VAL_3042]], float* %[[VAL_5]] |
| // CHECK: %[[VAL_3045:.*]] = load float, float* %[[VAL_3044]], align 4 |
| // CHECK: %[[VAL_3046:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3045]], i32 16, i32 31) |
| // CHECK: store float %[[VAL_3046]], float* %[[VAL_4]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_3044]], float* %[[VAL_4]], float* %[[VAL_3044]]) |
| // CHECK: %[[VAL_3047:.*]] = load float, float* %[[VAL_3044]], align 4 |
| // CHECK: %[[VAL_3048:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3047]], i32 8, i32 31) |
| // CHECK: store float %[[VAL_3048]], float* %[[VAL_3]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_3044]], float* %[[VAL_3]], float* %[[VAL_3044]]) |
| // CHECK: %[[VAL_3049:.*]] = load float, float* %[[VAL_3044]], align 4 |
| // CHECK: %[[VAL_3050:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3049]], i32 4, i32 31) |
| // CHECK: store float %[[VAL_3050]], float* %[[VAL_2]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_3044]], float* %[[VAL_2]], float* %[[VAL_3044]]) |
| // CHECK: %[[VAL_3051:.*]] = load float, float* %[[VAL_3044]], align 4 |
| // CHECK: %[[VAL_3052:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3051]], i32 2, i32 31) |
| // CHECK: store float %[[VAL_3052]], float* %[[VAL_1]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_3044]], float* %[[VAL_1]], float* %[[VAL_3044]]) |
| // CHECK: %[[VAL_3053:.*]] = load float, float* %[[VAL_3044]], align 4 |
| // CHECK: %[[VAL_3054:.*]] = call float @llvm.nvvm.shfl.sync.down.f32(i32 -1, float %[[VAL_3053]], i32 1, i32 31) |
| // CHECK: store float %[[VAL_3054]], float* %[[VAL_0]], align 4 |
| // CHECK: call void @region_2_9(float* %[[VAL_3044]], float* %[[VAL_0]], float* %[[VAL_3044]]) |
| // CHECK: %[[VAL_3055:.*]] = icmp eq i32 %[[VAL_84]], 0 |
| // CHECK: br i1 %[[VAL_3055]], label %[[VAL_3056:.*]], label %[[VAL_136]] |
| // CHECK: reduction_write_output-after816: ; preds = %[[VAL_3056]], %[[VAL_135]] |
| // CHECK: br label %[[VAL_46]] |
| // CHECK: reduction_write_output-true815: ; preds = %[[VAL_135]] |
| // CHECK: %[[VAL_3057:.*]] = load float, float* %[[VAL_3042]], align 4 |
| // CHECK: store float %[[VAL_3057]], float* %[[VAL_118]], align 4 |
| // CHECK: br label %[[VAL_136]] |
| // CHECK: entry: |
| // CHECK: %[[VAL_3058:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_3059:.*]] = load float, float* %[[VAL_3060:.*]], align 4 |
| // CHECK: %[[VAL_3061:.*]] = load float, float* %[[VAL_3062:.*]], align 4 |
| // CHECK: %[[VAL_3063:.*]] = fadd float %[[VAL_3059]], %[[VAL_3061]] |
| // CHECK: store float %[[VAL_3063]], float* %[[VAL_3058]], align 4 |
| // CHECK: %[[VAL_3064:.*]] = load float, float* %[[VAL_3058]], align 4 |
| // CHECK: store float %[[VAL_3064]], float* %[[VAL_3065:.*]], align 4 |
| // CHECK: ret void |
| // CHECK: entry: |
| // CHECK: %[[VAL_3066:.*]] = alloca float, align 4 |
| // CHECK: %[[VAL_3067:.*]] = load float, float* %[[VAL_3068:.*]], align 4 |
| // CHECK: %[[VAL_3069:.*]] = load float, float* %[[VAL_3070:.*]], align 4 |
| // CHECK: %[[VAL_3071:.*]] = call float @llvm.maxnum.f32(float %[[VAL_3067]], float %[[VAL_3069]]) |
| // CHECK: store float %[[VAL_3071]], float* %[[VAL_3066]], align 4 |
| // CHECK: %[[VAL_3072:.*]] = load float, float* %[[VAL_3066]], align 4 |
| // CHECK: store float %[[VAL_3072]], float* %[[VAL_3073:.*]], align 4 |
| // CHECK: ret void |
| |
| HloModule Test |
| |
| Add { |
| lhsadd = f32[] parameter(0) |
| rhsadd = f32[] parameter(1) |
| ROOT add = f32[] add(lhsadd, rhsadd) |
| } |
| |
| Max { |
| lhsmax = f32[] parameter(0) |
| rhsmax = f32[] parameter(1) |
| ROOT max = f32[] maximum(lhsmax, rhsmax) |
| } |
| |
| |
| fused_reduce { |
| p0 = f32[2,32,32]{2,1,0} parameter(0) |
| init1 = f32[] parameter(1) |
| init2 = f32[] parameter(2) |
| r1 = f32[2,32]{1,0} reduce(p0, init1), dimensions={2}, to_apply=Add |
| r2 = f32[2,32]{1,0} reduce(p0, init2), dimensions={2}, to_apply=Max |
| ROOT tuple = (f32[2,32]{1,0}, f32[2,32]{1,0}) tuple(r1, r2) |
| } |
| |
| ENTRY reduce { |
| p = f32[2,32,32]{2,1,0} parameter(0) |
| i = f32[] parameter(1) |
| j = f32[] parameter(2) |
| ROOT fusion = (f32[2,32]{1,0}, f32[2,32]{1,0}) fusion(p, i, j), |
| kind=kInput, calls=fused_reduce |
| } |
