Skip to content

[Loads] Check context instruction for context-sensitive derefability #109277

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
merged 2 commits into from
Sep 23, 2024
Merged

[Loads] Check context instruction for context-sensitive derefability #109277

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
edbdc03
[Loads] Check context instruction for context-sensitive derefability
nikic Sep 17, 2024
f4a78d4
add comment
nikic Sep 20, 2024
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
6 changes: 1 addition & 5 deletions clang/test/CodeGenOpenCL/builtins-amdgcn.cl
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -638,11 +638,7 @@ void test_get_workgroup_size(int d, global int *out)

// CHECK-LABEL: @test_get_grid_size(
// CHECK: {{.*}}call align 4 dereferenceable(64){{.*}} ptr addrspace(4) @llvm.amdgcn.dispatch.ptr()
// CHECK: getelementptr inbounds i8, ptr addrspace(4) %{{.*}}, i64 12
// CHECK: load i32, ptr addrspace(4) %{{.*}}, align 4, !invariant.load
// CHECK: getelementptr inbounds i8, ptr addrspace(4) %{{.*}}, i64 16
// CHECK: load i32, ptr addrspace(4) %{{.*}}, align 4, !invariant.load
// CHECK: getelementptr inbounds i8, ptr addrspace(4) %{{.*}}, i64 20
// CHECK: getelementptr inbounds i8, ptr addrspace(4) %{{.*}}, i64 %.sink
// CHECK: load i32, ptr addrspace(4) %{{.*}}, align 4, !invariant.load
void test_get_grid_size(int d, global int *out)
{
Expand Down
4 changes: 3 additions & 1 deletion llvm/include/llvm/Analysis/ValueTracking.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -805,7 +805,9 @@ bool onlyUsedByLifetimeMarkersOrDroppableInsts(const Value *V);
///
/// If the CtxI is specified this method performs context-sensitive analysis
/// and returns true if it is safe to execute the instruction immediately
/// before the CtxI.
/// before the CtxI. If the instruction has (transitive) operands that don't
/// dominate CtxI, the analysis is performed under the assumption that these
/// operands will also be speculated to a point before CxtI.
///
/// If the CtxI is NOT specified this method only looks at the instruction
/// itself and its operands, so if this method returns true, it is safe to
Expand Down
11 changes: 11 additions & 0 deletions llvm/lib/Analysis/Loads.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -104,6 +104,17 @@ static bool isDereferenceableAndAlignedPointer(
if (CheckForNonNull &&
!isKnownNonZero(V, SimplifyQuery(DL, DT, AC, CtxI)))
return false;
// When using something like !dereferenceable on a load, the
// dereferenceability may only be valid on a specific control-flow path.
// If the instruction doesn't dominate the context instruction, we're
// asking about dereferenceability under the assumption that the
// instruction has been speculated to the point of the context instruction,
// in which case we don't know if the dereferenceability info still holds.
// We don't bother handling allocas here, as they aren't speculatable
// anyway.
auto *I = dyn_cast<Instruction>(V);
Copy link

@tschuett tschuett Sep 19, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more. 

if (auto *I = dyn_cast<Instruction>(V))
  if (!isa<AllocaInst>(I))
     return CtxI && isValidAssumeForContext(I, CtxI, DT);

Copy link

@tschuett tschuett Sep 19, 2024

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more. 

if (auto *I = dyn_cast<Instruction>(V))
  if (!isa<AllocaInst>(I))
    if (CtxI == nullptr)
      return false;
    else
     return isValidAssumeForContext(I, CtxI, DT);

if (I && !isa<AllocaInst>(I))
return CtxI && isValidAssumeForContext(I, CtxI, DT);
return true;
};
if (IsKnownDeref()) {
Expand Down
4 changes: 2 additions & 2 deletions llvm/lib/Analysis/MemDerefPrinter.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -30,10 +30,10 @@ PreservedAnalyses MemDerefPrinterPass::run(Function &F,
for (auto &I : instructions(F)) {
if (LoadInst *LI = dyn_cast<LoadInst>(&I)) {
Value *PO = LI->getPointerOperand();
if (isDereferenceablePointer(PO, LI->getType(), DL))
if (isDereferenceablePointer(PO, LI->getType(), DL, LI))
Deref.push_back(PO);
if (isDereferenceableAndAlignedPointer(PO, LI->getType(), LI->getAlign(),
DL))
DL, LI))
DerefAndAligned.insert(PO);
}
}
Expand Down
3 changes: 2 additions & 1 deletion llvm/lib/CodeGen/MachineOperand.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -1047,7 +1047,8 @@ bool MachinePointerInfo::isDereferenceable(unsigned Size, LLVMContext &C,
return false;

return isDereferenceableAndAlignedPointer(
BasePtr, Align(1), APInt(DL.getPointerSizeInBits(), Offset + Size), DL);
BasePtr, Align(1), APInt(DL.getPointerSizeInBits(), Offset + Size), DL,
dyn_cast<Instruction>(BasePtr));
}

/// getConstantPool - Return a MachinePointerInfo record that refers to the
Expand Down
11 changes: 7 additions & 4 deletions llvm/test/Transforms/SimplifyCFG/speculate-derefable-load.ll
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -77,14 +77,17 @@ exit:
ret i64 %res
}

; FIXME: This is a miscompile.
define i64 @deref_no_hoist(i1 %c, ptr align 8 dereferenceable(8) %p1) {
; CHECK-LABEL: define i64 @deref_no_hoist(
; CHECK-SAME: i1 [[C:%.*]], ptr align 8 dereferenceable(8) [[P1:%.*]]) {
; CHECK-NEXT: [[ENTRY:.*:]]
; CHECK-NEXT: [[P2:%.*]] = load ptr, ptr [[P1]], align 8, !align [[META0:![0-9]+]]
; CHECK-NEXT: [[ENTRY:.*]]:
; CHECK-NEXT: br i1 [[C]], label %[[IF:.*]], label %[[EXIT:.*]]
; CHECK: [[IF]]:
; CHECK-NEXT: [[P2:%.*]] = load ptr, ptr [[P1]], align 8, !dereferenceable [[META0:![0-9]+]], !align [[META0]]
; CHECK-NEXT: [[V:%.*]] = load i64, ptr [[P2]], align 8
; CHECK-NEXT: [[RES:%.*]] = select i1 [[C]], i64 [[V]], i64 0
; CHECK-NEXT: br label %[[EXIT]]
; CHECK: [[EXIT]]:
; CHECK-NEXT: [[RES:%.*]] = phi i64 [ [[V]], %[[IF]] ], [ 0, %[[ENTRY]] ]
; CHECK-NEXT: ret i64 [[RES]]
;
entry:
Expand Down
Loading