llvmbot wrote:
<!--LLVM PR SUMMARY COMMENT--> @llvm/pr-subscribers-llvm-transforms Author: None (LU-JOHN) <details> <summary>Changes</summary> Extend jump-threading to allow local defs that are live outside of the threaded block. Allow threading to destinations where the local defs are not live. --- Patch is 127.83 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/135079.diff 5 Files Affected: - (modified) clang/test/CodeGenObjC/exceptions.m (+9-10) - (modified) clang/test/Headers/__clang_hip_math.hip (+525-581) - (modified) llvm/lib/Transforms/Utils/SimplifyCFG.cpp (+42-8) - (modified) llvm/test/CodeGen/AArch64/avoid-free-ext-promotion.ll (+16-20) - (added) llvm/test/Transforms/SimplifyCFG/jump-threading-live-on-exit.ll (+195) ``````````diff diff --git a/clang/test/CodeGenObjC/exceptions.m b/clang/test/CodeGenObjC/exceptions.m index 1546ed2585db5..832d3a458050c 100644 --- a/clang/test/CodeGenObjC/exceptions.m +++ b/clang/test/CodeGenObjC/exceptions.m @@ -144,18 +144,17 @@ void f4(void) { // CHECK-NEXT: br label // -> rethrow - // finally.call-exit: Predecessors are the @try and @catch fallthroughs - // as well as the no-match case in the catch mechanism. The i1 is whether - // to rethrow and should be true only in the last case. - // CHECK: phi ptr - // CHECK-NEXT: phi i1 - // CHECK-NEXT: call void @objc_exception_try_exit(ptr nonnull [[EXNDATA]]) + // finally.call-exit: Predecessor is the no-match case in the catch mechanism + // which rethrows. + // CHECK: call void @objc_exception_try_exit(ptr nonnull [[EXNDATA]]) // CHECK-NEXT: call void @f4_help(i32 noundef 2) - // CHECK-NEXT: br i1 - // -> ret, rethrow + // CHECK-NEXT: br label + // -> rethrow - // ret: - // CHECK: ret void + // finally.end.critedge: Predecessors are the @try and @catch fallthroughs. + // CHECK: call void @objc_exception_try_exit(ptr nonnull [[EXNDATA]]) + // CHECK-NEXT: call void @f4_help(i32 noundef 2) + // CHECK-NEXT: ret void // Catch mechanism: // CHECK: call ptr @objc_exception_extract(ptr nonnull [[EXNDATA]]) diff --git a/clang/test/Headers/__clang_hip_math.hip b/clang/test/Headers/__clang_hip_math.hip index e879fec0ebe5a..dba17478ecca9 100644 --- a/clang/test/Headers/__clang_hip_math.hip +++ b/clang/test/Headers/__clang_hip_math.hip @@ -40,30 +40,27 @@ typedef unsigned long long uint64_t; // CHECK-LABEL: @test___make_mantissa_base8( // CHECK-NEXT: entry: -// CHECK-NEXT: br label [[WHILE_COND_I:%.*]] -// CHECK: while.cond.i: -// CHECK-NEXT: [[__TAGP_ADDR_0_I:%.*]] = phi ptr [ [[P:%.*]], [[ENTRY:%.*]] ], [ [[__TAGP_ADDR_1_I:%.*]], [[CLEANUP_I:%.*]] ] -// CHECK-NEXT: [[__R_0_I:%.*]] = phi i64 [ 0, [[ENTRY]] ], [ [[__R_1_I:%.*]], [[CLEANUP_I]] ] -// CHECK-NEXT: [[TMP0:%.*]] = load i8, ptr [[__TAGP_ADDR_0_I]], align 1, !tbaa [[TBAA4:![0-9]+]] -// CHECK-NEXT: [[CMP_NOT_I:%.*]] = icmp eq i8 [[TMP0]], 0 -// CHECK-NEXT: br i1 [[CMP_NOT_I]], label [[_ZL21__MAKE_MANTISSA_BASE8PKC_EXIT:%.*]], label [[WHILE_BODY_I:%.*]] +// CHECK-NEXT: [[TMP0:%.*]] = load i8, ptr [[P:%.*]], align 1, !tbaa [[TBAA4:![0-9]+]] +// CHECK-NEXT: [[CMP_NOT_I1:%.*]] = icmp eq i8 [[TMP0]], 0 +// CHECK-NEXT: br i1 [[CMP_NOT_I1]], label [[_ZL21__MAKE_MANTISSA_BASE8PKC_EXIT:%.*]], label [[WHILE_BODY_I:%.*]] // CHECK: while.body.i: -// CHECK-NEXT: [[TMP1:%.*]] = and i8 [[TMP0]], -8 -// CHECK-NEXT: [[OR_COND_I:%.*]] = icmp eq i8 [[TMP1]], 48 -// CHECK-NEXT: br i1 [[OR_COND_I]], label [[IF_THEN_I:%.*]], label [[CLEANUP_I]] +// CHECK-NEXT: [[TMP1:%.*]] = phi i8 [ [[TMP3:%.*]], [[IF_THEN_I:%.*]] ], [ [[TMP0]], [[ENTRY:%.*]] ] +// CHECK-NEXT: [[__R_0_I3:%.*]] = phi i64 [ [[SUB_I:%.*]], [[IF_THEN_I]] ], [ 0, [[ENTRY]] ] +// CHECK-NEXT: [[__TAGP_ADDR_0_I2:%.*]] = phi ptr [ [[INCDEC_PTR_I:%.*]], [[IF_THEN_I]] ], [ [[P]], [[ENTRY]] ] +// CHECK-NEXT: [[TMP2:%.*]] = and i8 [[TMP1]], -8 +// CHECK-NEXT: [[OR_COND_I:%.*]] = icmp eq i8 [[TMP2]], 48 +// CHECK-NEXT: br i1 [[OR_COND_I]], label [[IF_THEN_I]], label [[_ZL21__MAKE_MANTISSA_BASE8PKC_EXIT]] // CHECK: if.then.i: -// CHECK-NEXT: [[MUL_I:%.*]] = shl i64 [[__R_0_I]], 3 -// CHECK-NEXT: [[CONV5_I:%.*]] = zext nneg i8 [[TMP0]] to i64 +// CHECK-NEXT: [[MUL_I:%.*]] = shl i64 [[__R_0_I3]], 3 +// CHECK-NEXT: [[CONV5_I:%.*]] = zext nneg i8 [[TMP1]] to i64 // CHECK-NEXT: [[ADD_I:%.*]] = add i64 [[MUL_I]], -48 -// CHECK-NEXT: [[SUB_I:%.*]] = add i64 [[ADD_I]], [[CONV5_I]] -// CHECK-NEXT: [[INCDEC_PTR_I:%.*]] = getelementptr inbounds nuw i8, ptr [[__TAGP_ADDR_0_I]], i64 1 -// CHECK-NEXT: br label [[CLEANUP_I]] -// CHECK: cleanup.i: -// CHECK-NEXT: [[__TAGP_ADDR_1_I]] = phi ptr [ [[INCDEC_PTR_I]], [[IF_THEN_I]] ], [ [[__TAGP_ADDR_0_I]], [[WHILE_BODY_I]] ] -// CHECK-NEXT: [[__R_1_I]] = phi i64 [ [[SUB_I]], [[IF_THEN_I]] ], [ [[__R_0_I]], [[WHILE_BODY_I]] ] -// CHECK-NEXT: br i1 [[OR_COND_I]], label [[WHILE_COND_I]], label [[_ZL21__MAKE_MANTISSA_BASE8PKC_EXIT]], !llvm.loop [[LOOP7:![0-9]+]] +// CHECK-NEXT: [[SUB_I]] = add i64 [[ADD_I]], [[CONV5_I]] +// CHECK-NEXT: [[INCDEC_PTR_I]] = getelementptr inbounds nuw i8, ptr [[__TAGP_ADDR_0_I2]], i64 1 +// CHECK-NEXT: [[TMP3]] = load i8, ptr [[INCDEC_PTR_I]], align 1, !tbaa [[TBAA4]] +// CHECK-NEXT: [[CMP_NOT_I:%.*]] = icmp eq i8 [[TMP3]], 0 +// CHECK-NEXT: br i1 [[CMP_NOT_I]], label [[_ZL21__MAKE_MANTISSA_BASE8PKC_EXIT]], label [[WHILE_BODY_I]], !llvm.loop [[LOOP7:![0-9]+]] // CHECK: _ZL21__make_mantissa_base8PKc.exit: -// CHECK-NEXT: [[RETVAL_2_I:%.*]] = phi i64 [ 0, [[CLEANUP_I]] ], [ [[__R_0_I]], [[WHILE_COND_I]] ] +// CHECK-NEXT: [[RETVAL_2_I:%.*]] = phi i64 [ 0, [[ENTRY]] ], [ 0, [[WHILE_BODY_I]] ], [ [[SUB_I]], [[IF_THEN_I]] ] // CHECK-NEXT: ret i64 [[RETVAL_2_I]] // // AMDGCNSPIRV-LABEL: @test___make_mantissa_base8( @@ -96,30 +93,27 @@ extern "C" __device__ uint64_t test___make_mantissa_base8(const char *p) { // CHECK-LABEL: @test___make_mantissa_base10( // CHECK-NEXT: entry: -// CHECK-NEXT: br label [[WHILE_COND_I:%.*]] -// CHECK: while.cond.i: -// CHECK-NEXT: [[__TAGP_ADDR_0_I:%.*]] = phi ptr [ [[P:%.*]], [[ENTRY:%.*]] ], [ [[__TAGP_ADDR_1_I:%.*]], [[CLEANUP_I:%.*]] ] -// CHECK-NEXT: [[__R_0_I:%.*]] = phi i64 [ 0, [[ENTRY]] ], [ [[__R_1_I:%.*]], [[CLEANUP_I]] ] -// CHECK-NEXT: [[TMP0:%.*]] = load i8, ptr [[__TAGP_ADDR_0_I]], align 1, !tbaa [[TBAA4]] -// CHECK-NEXT: [[CMP_NOT_I:%.*]] = icmp eq i8 [[TMP0]], 0 -// CHECK-NEXT: br i1 [[CMP_NOT_I]], label [[_ZL22__MAKE_MANTISSA_BASE10PKC_EXIT:%.*]], label [[WHILE_BODY_I:%.*]] +// CHECK-NEXT: [[TMP0:%.*]] = load i8, ptr [[P:%.*]], align 1, !tbaa [[TBAA4]] +// CHECK-NEXT: [[CMP_NOT_I1:%.*]] = icmp eq i8 [[TMP0]], 0 +// CHECK-NEXT: br i1 [[CMP_NOT_I1]], label [[_ZL22__MAKE_MANTISSA_BASE10PKC_EXIT:%.*]], label [[WHILE_BODY_I:%.*]] // CHECK: while.body.i: -// CHECK-NEXT: [[TMP1:%.*]] = add i8 [[TMP0]], -48 -// CHECK-NEXT: [[OR_COND_I:%.*]] = icmp ult i8 [[TMP1]], 10 -// CHECK-NEXT: br i1 [[OR_COND_I]], label [[IF_THEN_I:%.*]], label [[CLEANUP_I]] +// CHECK-NEXT: [[TMP1:%.*]] = phi i8 [ [[TMP3:%.*]], [[IF_THEN_I:%.*]] ], [ [[TMP0]], [[ENTRY:%.*]] ] +// CHECK-NEXT: [[__R_0_I3:%.*]] = phi i64 [ [[SUB_I:%.*]], [[IF_THEN_I]] ], [ 0, [[ENTRY]] ] +// CHECK-NEXT: [[__TAGP_ADDR_0_I2:%.*]] = phi ptr [ [[INCDEC_PTR_I:%.*]], [[IF_THEN_I]] ], [ [[P]], [[ENTRY]] ] +// CHECK-NEXT: [[TMP2:%.*]] = add i8 [[TMP1]], -48 +// CHECK-NEXT: [[OR_COND_I:%.*]] = icmp ult i8 [[TMP2]], 10 +// CHECK-NEXT: br i1 [[OR_COND_I]], label [[IF_THEN_I]], label [[_ZL22__MAKE_MANTISSA_BASE10PKC_EXIT]] // CHECK: if.then.i: -// CHECK-NEXT: [[MUL_I:%.*]] = mul i64 [[__R_0_I]], 10 -// CHECK-NEXT: [[CONV5_I:%.*]] = zext nneg i8 [[TMP0]] to i64 +// CHECK-NEXT: [[MUL_I:%.*]] = mul i64 [[__R_0_I3]], 10 +// CHECK-NEXT: [[CONV5_I:%.*]] = zext nneg i8 [[TMP1]] to i64 // CHECK-NEXT: [[ADD_I:%.*]] = add i64 [[MUL_I]], -48 -// CHECK-NEXT: [[SUB_I:%.*]] = add i64 [[ADD_I]], [[CONV5_I]] -// CHECK-NEXT: [[INCDEC_PTR_I:%.*]] = getelementptr inbounds nuw i8, ptr [[__TAGP_ADDR_0_I]], i64 1 -// CHECK-NEXT: br label [[CLEANUP_I]] -// CHECK: cleanup.i: -// CHECK-NEXT: [[__TAGP_ADDR_1_I]] = phi ptr [ [[INCDEC_PTR_I]], [[IF_THEN_I]] ], [ [[__TAGP_ADDR_0_I]], [[WHILE_BODY_I]] ] -// CHECK-NEXT: [[__R_1_I]] = phi i64 [ [[SUB_I]], [[IF_THEN_I]] ], [ [[__R_0_I]], [[WHILE_BODY_I]] ] -// CHECK-NEXT: br i1 [[OR_COND_I]], label [[WHILE_COND_I]], label [[_ZL22__MAKE_MANTISSA_BASE10PKC_EXIT]], !llvm.loop [[LOOP10:![0-9]+]] +// CHECK-NEXT: [[SUB_I]] = add i64 [[ADD_I]], [[CONV5_I]] +// CHECK-NEXT: [[INCDEC_PTR_I]] = getelementptr inbounds nuw i8, ptr [[__TAGP_ADDR_0_I2]], i64 1 +// CHECK-NEXT: [[TMP3]] = load i8, ptr [[INCDEC_PTR_I]], align 1, !tbaa [[TBAA4]] +// CHECK-NEXT: [[CMP_NOT_I:%.*]] = icmp eq i8 [[TMP3]], 0 +// CHECK-NEXT: br i1 [[CMP_NOT_I]], label [[_ZL22__MAKE_MANTISSA_BASE10PKC_EXIT]], label [[WHILE_BODY_I]], !llvm.loop [[LOOP10:![0-9]+]] // CHECK: _ZL22__make_mantissa_base10PKc.exit: -// CHECK-NEXT: [[RETVAL_2_I:%.*]] = phi i64 [ 0, [[CLEANUP_I]] ], [ [[__R_0_I]], [[WHILE_COND_I]] ] +// CHECK-NEXT: [[RETVAL_2_I:%.*]] = phi i64 [ 0, [[ENTRY]] ], [ 0, [[WHILE_BODY_I]] ], [ [[SUB_I]], [[IF_THEN_I]] ] // CHECK-NEXT: ret i64 [[RETVAL_2_I]] // // AMDGCNSPIRV-LABEL: @test___make_mantissa_base10( @@ -152,78 +146,70 @@ extern "C" __device__ uint64_t test___make_mantissa_base10(const char *p) { // CHECK-LABEL: @test___make_mantissa_base16( // CHECK-NEXT: entry: -// CHECK-NEXT: br label [[WHILE_COND_I:%.*]] -// CHECK: while.cond.i: -// CHECK-NEXT: [[__TAGP_ADDR_0_I:%.*]] = phi ptr [ [[P:%.*]], [[ENTRY:%.*]] ], [ [[__TAGP_ADDR_1_I:%.*]], [[CLEANUP_I:%.*]] ] -// CHECK-NEXT: [[__R_0_I:%.*]] = phi i64 [ 0, [[ENTRY]] ], [ [[__R_2_I:%.*]], [[CLEANUP_I]] ] -// CHECK-NEXT: [[TMP0:%.*]] = load i8, ptr [[__TAGP_ADDR_0_I]], align 1, !tbaa [[TBAA4]] -// CHECK-NEXT: [[CMP_NOT_I:%.*]] = icmp eq i8 [[TMP0]], 0 -// CHECK-NEXT: br i1 [[CMP_NOT_I]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT:%.*]], label [[WHILE_BODY_I:%.*]] +// CHECK-NEXT: [[TMP0:%.*]] = load i8, ptr [[P:%.*]], align 1, !tbaa [[TBAA4]] +// CHECK-NEXT: [[CMP_NOT_I1:%.*]] = icmp eq i8 [[TMP0]], 0 +// CHECK-NEXT: br i1 [[CMP_NOT_I1]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT:%.*]], label [[WHILE_BODY_I:%.*]] // CHECK: while.body.i: -// CHECK-NEXT: [[TMP1:%.*]] = add i8 [[TMP0]], -48 -// CHECK-NEXT: [[OR_COND_I:%.*]] = icmp ult i8 [[TMP1]], 10 -// CHECK-NEXT: br i1 [[OR_COND_I]], label [[IF_END31_I:%.*]], label [[IF_ELSE_I:%.*]] +// CHECK-NEXT: [[TMP1:%.*]] = phi i8 [ [[TMP5:%.*]], [[IF_END31_I:%.*]] ], [ [[TMP0]], [[ENTRY:%.*]] ] +// CHECK-NEXT: [[__R_0_I3:%.*]] = phi i64 [ [[ADD28_I:%.*]], [[IF_END31_I]] ], [ 0, [[ENTRY]] ] +// CHECK-NEXT: [[__TAGP_ADDR_0_I2:%.*]] = phi ptr [ [[INCDEC_PTR_I:%.*]], [[IF_END31_I]] ], [ [[P]], [[ENTRY]] ] +// CHECK-NEXT: [[TMP2:%.*]] = add i8 [[TMP1]], -48 +// CHECK-NEXT: [[OR_COND_I:%.*]] = icmp ult i8 [[TMP2]], 10 +// CHECK-NEXT: br i1 [[OR_COND_I]], label [[IF_END31_I]], label [[IF_ELSE_I:%.*]] // CHECK: if.else.i: -// CHECK-NEXT: [[TMP2:%.*]] = add i8 [[TMP0]], -97 -// CHECK-NEXT: [[OR_COND33_I:%.*]] = icmp ult i8 [[TMP2]], 6 +// CHECK-NEXT: [[TMP3:%.*]] = add i8 [[TMP1]], -97 +// CHECK-NEXT: [[OR_COND33_I:%.*]] = icmp ult i8 [[TMP3]], 6 // CHECK-NEXT: br i1 [[OR_COND33_I]], label [[IF_END31_I]], label [[IF_ELSE17_I:%.*]] // CHECK: if.else17.i: -// CHECK-NEXT: [[TMP3:%.*]] = add i8 [[TMP0]], -65 -// CHECK-NEXT: [[OR_COND34_I:%.*]] = icmp ult i8 [[TMP3]], 6 -// CHECK-NEXT: br i1 [[OR_COND34_I]], label [[IF_END31_I]], label [[CLEANUP_I]] +// CHECK-NEXT: [[TMP4:%.*]] = add i8 [[TMP1]], -65 +// CHECK-NEXT: [[OR_COND34_I:%.*]] = icmp ult i8 [[TMP4]], 6 +// CHECK-NEXT: br i1 [[OR_COND34_I]], label [[IF_END31_I]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT]] // CHECK: if.end31.i: // CHECK-NEXT: [[DOTSINK:%.*]] = phi i64 [ -48, [[WHILE_BODY_I]] ], [ -87, [[IF_ELSE_I]] ], [ -55, [[IF_ELSE17_I]] ] -// CHECK-NEXT: [[MUL24_I:%.*]] = shl i64 [[__R_0_I]], 4 -// CHECK-NEXT: [[CONV25_I:%.*]] = zext nneg i8 [[TMP0]] to i64 +// CHECK-NEXT: [[MUL24_I:%.*]] = shl i64 [[__R_0_I3]], 4 +// CHECK-NEXT: [[CONV25_I:%.*]] = zext nneg i8 [[TMP1]] to i64 // CHECK-NEXT: [[ADD26_I:%.*]] = add i64 [[MUL24_I]], [[DOTSINK]] -// CHECK-NEXT: [[ADD28_I:%.*]] = add i64 [[ADD26_I]], [[CONV25_I]] -// CHECK-NEXT: [[INCDEC_PTR_I:%.*]] = getelementptr inbounds nuw i8, ptr [[__TAGP_ADDR_0_I]], i64 1 -// CHECK-NEXT: br label [[CLEANUP_I]] -// CHECK: cleanup.i: -// CHECK-NEXT: [[__TAGP_ADDR_1_I]] = phi ptr [ [[INCDEC_PTR_I]], [[IF_END31_I]] ], [ [[__TAGP_ADDR_0_I]], [[IF_ELSE17_I]] ] -// CHECK-NEXT: [[__R_2_I]] = phi i64 [ [[ADD28_I]], [[IF_END31_I]] ], [ [[__R_0_I]], [[IF_ELSE17_I]] ] -// CHECK-NEXT: [[COND_I:%.*]] = phi i1 [ true, [[IF_END31_I]] ], [ false, [[IF_ELSE17_I]] ] -// CHECK-NEXT: br i1 [[COND_I]], label [[WHILE_COND_I]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT]], !llvm.loop [[LOOP11:![0-9]+]] +// CHECK-NEXT: [[ADD28_I]] = add i64 [[ADD26_I]], [[CONV25_I]] +// CHECK-NEXT: [[INCDEC_PTR_I]] = getelementptr inbounds nuw i8, ptr [[__TAGP_ADDR_0_I2]], i64 1 +// CHECK-NEXT: [[TMP5]] = load i8, ptr [[INCDEC_PTR_I]], align 1, !tbaa [[TBAA4]] +// CHECK-NEXT: [[CMP_NOT_I:%.*]] = icmp eq i8 [[TMP5]], 0 +// CHECK-NEXT: br i1 [[CMP_NOT_I]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT]], label [[WHILE_BODY_I]], !llvm.loop [[LOOP11:![0-9]+]] // CHECK: _ZL22__make_mantissa_base16PKc.exit: -// CHECK-NEXT: [[RETVAL_2_I:%.*]] = phi i64 [ 0, [[CLEANUP_I]] ], [ [[__R_0_I]], [[WHILE_COND_I]] ] +// CHECK-NEXT: [[RETVAL_2_I:%.*]] = phi i64 [ 0, [[ENTRY]] ], [ 0, [[IF_ELSE17_I]] ], [ [[ADD28_I]], [[IF_END31_I]] ] // CHECK-NEXT: ret i64 [[RETVAL_2_I]] // // AMDGCNSPIRV-LABEL: @test___make_mantissa_base16( // AMDGCNSPIRV-NEXT: entry: -// AMDGCNSPIRV-NEXT: br label [[WHILE_COND_I:%.*]] -// AMDGCNSPIRV: while.cond.i: -// AMDGCNSPIRV-NEXT: [[__TAGP_ADDR_0_I:%.*]] = phi ptr addrspace(4) [ [[P:%.*]], [[ENTRY:%.*]] ], [ [[__TAGP_ADDR_1_I:%.*]], [[CLEANUP_I:%.*]] ] -// AMDGCNSPIRV-NEXT: [[__R_0_I:%.*]] = phi i64 [ 0, [[ENTRY]] ], [ [[__R_2_I:%.*]], [[CLEANUP_I]] ] -// AMDGCNSPIRV-NEXT: [[TMP0:%.*]] = load i8, ptr addrspace(4) [[__TAGP_ADDR_0_I]], align 1, !tbaa [[TBAA5]] -// AMDGCNSPIRV-NEXT: [[CMP_NOT_I:%.*]] = icmp eq i8 [[TMP0]], 0 -// AMDGCNSPIRV-NEXT: br i1 [[CMP_NOT_I]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT:%.*]], label [[WHILE_BODY_I:%.*]] +// AMDGCNSPIRV-NEXT: [[TMP0:%.*]] = load i8, ptr addrspace(4) [[P:%.*]], align 1, !tbaa [[TBAA5]] +// AMDGCNSPIRV-NEXT: [[CMP_NOT_I1:%.*]] = icmp eq i8 [[TMP0]], 0 +// AMDGCNSPIRV-NEXT: br i1 [[CMP_NOT_I1]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT:%.*]], label [[WHILE_BODY_I:%.*]] // AMDGCNSPIRV: while.body.i: -// AMDGCNSPIRV-NEXT: [[TMP1:%.*]] = add i8 [[TMP0]], -48 -// AMDGCNSPIRV-NEXT: [[OR_COND_I:%.*]] = icmp ult i8 [[TMP1]], 10 -// AMDGCNSPIRV-NEXT: br i1 [[OR_COND_I]], label [[IF_END31_I:%.*]], label [[IF_ELSE_I:%.*]] +// AMDGCNSPIRV-NEXT: [[TMP1:%.*]] = phi i8 [ [[TMP5:%.*]], [[IF_END31_I:%.*]] ], [ [[TMP0]], [[ENTRY:%.*]] ] +// AMDGCNSPIRV-NEXT: [[__R_0_I3:%.*]] = phi i64 [ [[ADD28_I:%.*]], [[IF_END31_I]] ], [ 0, [[ENTRY]] ] +// AMDGCNSPIRV-NEXT: [[__TAGP_ADDR_0_I2:%.*]] = phi ptr addrspace(4) [ [[INCDEC_PTR_I:%.*]], [[IF_END31_I]] ], [ [[P]], [[ENTRY]] ] +// AMDGCNSPIRV-NEXT: [[TMP2:%.*]] = add i8 [[TMP1]], -48 +// AMDGCNSPIRV-NEXT: [[OR_COND_I:%.*]] = icmp ult i8 [[TMP2]], 10 +// AMDGCNSPIRV-NEXT: br i1 [[OR_COND_I]], label [[IF_END31_I]], label [[IF_ELSE_I:%.*]] // AMDGCNSPIRV: if.else.i: -// AMDGCNSPIRV-NEXT: [[TMP2:%.*]] = add i8 [[TMP0]], -97 -// AMDGCNSPIRV-NEXT: [[OR_COND33_I:%.*]] = icmp ult i8 [[TMP2]], 6 +// AMDGCNSPIRV-NEXT: [[TMP3:%.*]] = add i8 [[TMP1]], -97 +// AMDGCNSPIRV-NEXT: [[OR_COND33_I:%.*]] = icmp ult i8 [[TMP3]], 6 // AMDGCNSPIRV-NEXT: br i1 [[OR_COND33_I]], label [[IF_END31_I]], label [[IF_ELSE17_I:%.*]] // AMDGCNSPIRV: if.else17.i: -// AMDGCNSPIRV-NEXT: [[TMP3:%.*]] = add i8 [[TMP0]], -65 -// AMDGCNSPIRV-NEXT: [[OR_COND34_I:%.*]] = icmp ult i8 [[TMP3]], 6 -// AMDGCNSPIRV-NEXT: br i1 [[OR_COND34_I]], label [[IF_END31_I]], label [[CLEANUP_I]] +// AMDGCNSPIRV-NEXT: [[TMP4:%.*]] = add i8 [[TMP1]], -65 +// AMDGCNSPIRV-NEXT: [[OR_COND34_I:%.*]] = icmp ult i8 [[TMP4]], 6 +// AMDGCNSPIRV-NEXT: br i1 [[OR_COND34_I]], label [[IF_END31_I]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT]] // AMDGCNSPIRV: if.end31.i: // AMDGCNSPIRV-NEXT: [[DOTSINK:%.*]] = phi i64 [ -48, [[WHILE_BODY_I]] ], [ -87, [[IF_ELSE_I]] ], [ -55, [[IF_ELSE17_I]] ] -// AMDGCNSPIRV-NEXT: [[MUL24_I:%.*]] = shl i64 [[__R_0_I]], 4 -// AMDGCNSPIRV-NEXT: [[CONV25_I:%.*]] = zext nneg i8 [[TMP0]] to i64 +// AMDGCNSPIRV-NEXT: [[MUL24_I:%.*]] = shl i64 [[__R_0_I3]], 4 +// AMDGCNSPIRV-NEXT: [[CONV25_I:%.*]] = zext nneg i8 [[TMP1]] to i64 // AMDGCNSPIRV-NEXT: [[ADD26_I:%.*]] = add i64 [[MUL24_I]], [[DOTSINK]] -// AMDGCNSPIRV-NEXT: [[ADD28_I:%.*]] = add i64 [[ADD26_I]], [[CONV25_I]] -// AMDGCNSPIRV-NEXT: [[INCDEC_PTR_I:%.*]] = getelementptr inbounds nuw i8, ptr addrspace(4) [[__TAGP_ADDR_0_I]], i64 1 -// AMDGCNSPIRV-NEXT: br label [[CLEANUP_I]] -// AMDGCNSPIRV: cleanup.i: -// AMDGCNSPIRV-NEXT: [[__TAGP_ADDR_1_I]] = phi ptr addrspace(4) [ [[INCDEC_PTR_I]], [[IF_END31_I]] ], [ [[__TAGP_ADDR_0_I]], [[IF_ELSE17_I]] ] -// AMDGCNSPIRV-NEXT: [[__R_2_I]] = phi i64 [ [[ADD28_I]], [[IF_END31_I]] ], [ [[__R_0_I]], [[IF_ELSE17_I]] ] -// AMDGCNSPIRV-NEXT: [[COND_I:%.*]] = phi i1 [ true, [[IF_END31_I]] ], [ false, [[IF_ELSE17_I]] ] -// AMDGCNSPIRV-NEXT: br i1 [[COND_I]], label [[WHILE_COND_I]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT]], !llvm.loop [[LOOP12:![0-9]+]] +// AMDGCNSPIRV-NEXT: [[ADD28_I]] = add i64 [[ADD26_I]], [[CONV25_I]] +// AMDGCNSPIRV-NEXT: [[INCDEC_PTR_I]] = getelementptr inbounds nuw i8, ptr addrspace(4) [[__TAGP_ADDR_0_I2]], i64 1 +// AMDGCNSPIRV-NEXT: [[TMP5]] = load i8, ptr addrspace(4) [[INCDEC_PTR_I]], align 1, !tbaa [[TBAA5]] +// AMDGCNSPIRV-NEXT: [[CMP_NOT_I:%.*]] = icmp eq i8 [[TMP5]], 0 +// AMDGCNSPIRV-NEXT: br i1 [[CMP_NOT_I]], label [[_ZL22__MAKE_MANTISSA_BASE16PKC_EXIT]], label [[WHILE_BODY_I]], !llvm.loop [[LOOP12:![0-9]+]] // AMDGCNSPIRV: _ZL22__make_mantissa_base16PKc.exit: -// AMDGCNSPIRV-NEXT: [[RETVAL_2_I:%.*]] = phi i64 [ 0, [[CLEANUP_I]] ], [ [[__R_0_I]], [[WHILE_COND_I]] ] +// AMDGCNSPIRV-NEXT: [[RETVAL_2_I:%.*]] = phi i64 [ 0, [[ENTRY]] ], [ 0, [[IF_ELSE17_I]] ], [ [[ADD28_I]], [[IF_END31_I]] ] // AMDGCNSPIRV-NEXT: ret i64 [[RETVAL_2_I]] // extern "C" __device__ uint64_t test___make_mantissa_base16(const char *p) { @@ -234,91 +220,85 @@ extern "C" __device__ uint64_t test___make_mantissa_base16(const char *p) { // CHECK-NEXT: entry: // CHECK-NEXT: [[TMP0:%.*]] = load i8, ptr [[P:%.*]], align 1, !tbaa [[TBAA4]] // CHECK-NEXT: [[CMP_I:%.*]] = icmp eq i8 [[TMP0]], 48 -// CHECK-NEXT: br i1 [[CMP_I]], label [[IF_THEN_I:%.*]], label [[WHILE_COND_I14_I:%.*]] +// CHECK-NEXT: br i1 [[CMP_I]], label [[IF_THEN_I:%.*]], label [[WHILE_COND_I14_I_PREHEADER:%.*]] +// CHECK: while.cond.i14.i.preheader: +// CHECK-NEXT: [[TMP1:%.*]] = load i8, ptr [[P]], align 1, !tbaa [[TBAA4]] +// CHECK-NEXT: [[CMP_NOT_I17_I5:%.*]] = icmp eq i8 [[TMP1]], 0 +// CHECK-NEXT: br i1 [[CMP_NOT_I17_I5]], label [[_ZL15__MAKE_MANTISSAPKC_EXIT:%.*]], label [[WHILE_BODY_I18_I:%.*]] // CHECK: if.then.i: // CHECK-NEXT: [[INCDEC_PTR_I:%.*]] = getelementptr inbounds nuw i8, ptr [[P]], i64 1 -// CHECK-NEXT: [[TMP1:%.*]] = load i8, ptr [[INCDEC_PTR_I]], align 1, !tbaa [[TBAA4]] -// CHECK-NEXT: switch i8 [[TMP1]], label [[WHILE_COND_I_I:%.*]] [ -// CHECK-NEXT: i8 120, label [[WHILE_COND_I30_I_PREHEADER:%.*]] -// CHECK-NEXT: i8 88, label [[WHILE_COND_I30_I_PREHEADER]] +// CHECK-NEXT: [[TMP2:%.*]] = load i8, ptr [[INCDEC_PTR_I]], align 1, !tbaa [[TBAA4]] +// CHECK-NEXT: switch i8 [[TMP2]], label [[WHILE_COND_I_I_PREHEADER:%.*]] [ +// CHECK-NEXT: i8 120, label [[IF_THEN5_I:%.*]] +// CHECK-NEXT: i8 88, label [[IF_THEN5_I]] // CHECK-NEXT: ] -// CHECK: while.cond.i30.i.preheader: -// CHECK-NEXT: br label [[WHILE_COND_I30_I:%.*]] -// CHECK: while.cond.i30.i: -// CHECK-NEXT: [[__TAGP_ADDR_0_I31_I:%.*]] = phi ptr [ [[__TAGP_ADDR_1_I37_I:%.*]], [[CLEANUP_I36_I:%.*]] ], [ [[INCDEC_PTR_I]], [[WHILE_COND_I30_I_PREHEADER]] ] -// CHECK-NEXT: [[__R_0_I32_I:%.*]] = phi i64 [ [[__R_2_I_I:%.*]], [[CLEANUP_I36_I]] ], [ 0, [[WHILE_COND_I30_I_PREHEADER]] ] -// CHECK-NEXT: [[TMP2:%.*]] = load i8, ptr [[__TAGP_ADDR_0_I31_I]], align 1, !tbaa [[TBAA4]] -// CHECK-NEXT: [[CMP_NOT_I33_I:%.*]] = icmp eq i8 [[TMP2]], 0 -// CHECK-NEXT: br i1 [[CMP_NOT_I33_I]], label [[_ZL15__MAKE_MANTISSAPKC_EXIT:%.*]], label [[WHILE_BODY_I34_I:%.*]] -// CHECK: ... [truncated] `````````` </details> https://github.com/llvm/llvm-project/pull/135079 _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
