search for: valuetrack

...@llvm.xxx(), !range !0 !0 = metadata !{ i32 0, i23 1024 } The motivation behind this extension is some optimizations we are working on for CUDA programs. Some special registers in CUDA (e.g., threadIdx.x) are bounded per CUDA programming guide, and knowing their ranges can improve the precision of ValueTracking and benefit optimizations such as InstCombine. To implement this idea, we need ValueTracking to be aware of the ranges of these special variables. These special registers are so far read-only and accessed using intrinsics. e.g., %threadIdx.x = call i32 @llvm.nvvm.read.ptx.sreg.tid.x(). One pos...

Hello, As I can see, recently LLVM switched to using assume bundles to encode alignment information: https://reviews.llvm.org/rG78de7297abe2e8fa782682168989c70e3cb34a5c However, it seems that the ValueTracking cannot understand the new format. As an example, consider compilation of the following reproducer with clang-11 (old assume format) and clang-trunk (assume bundles): #include <stdint.h> int foo(int *p) { __builtin_assume_aligned(p, 32); return ((intptr_t) p) & 4; } clan...

...i32 0, i23 1024 } > > The motivation behind this extension is some optimizations we are > working on for CUDA programs. Some special registers in CUDA (e.g., > threadIdx.x) are bounded per CUDA programming guide, and knowing > their ranges can improve the precision of ValueTracking and benefit > optimizations such as InstCombine. > > To implement this idea, we need ValueTracking to be aware of the > ranges of these special variables. These special registers are so > far read-only and accessed using intrinsics. e.g., > %threadIdx.x = ca...

Some optimizations depend on whether alloca instruction always has non-zero value. Currently, this checking is done by isKnownNonZero() in ValueTracking, and it assumes alloca in address space 0 always has non-zero value but alloca in non-zero address spaces does not always have non-zero value. However, this assumption is incorrect for certain targets. For example, amdgcn---amdgiz target has alloca in address space 5, and its alloca always has...

...ics themselves >> into the passes if possible. Metadata will always be secondary. >> > > So you're saying that in this particular case you'd prefer LLVM passes to > know about the range of these PTX intrinsics, rather than Clang adding them > as metadata? > Yep. ValueTracking.cpp already has some iffy target knowledge (someone sneaked a > direct Intrinsic::x86_sse42_crc32_64_64 check in there), but extending it > to other intrinsics in other targets seems like too much... > That's not iffy. That's exactly how it should work, and we should have more...

...ways has non-zero value. Sam From: Hal Finkel [mailto:hfinkel at anl.gov] Sent: Thursday, December 14, 2017 4:06 PM To: Liu, Yaxun (Sam) <Yaxun.Liu at amd.com>; llvm-dev <llvm-dev at lists.llvm.org> Subject: Re: [llvm-dev] [RFC] Add TargetTransformInfo::isAllocaPtrValueNonZero and let ValueTracking depend on TargetTransformInfo Hi, Sam, Our general design has been that TTI is for cost modeling, not for providing semantic information. This piece of information can go in DataLayout, and I think we should put it there. Especially given that this affects our canonicalization process, or pre...

...econdary. >>> >>> >>> So you're saying that in this particular case you'd prefer LLVM passes to >>> know about the range of these PTX intrinsics, rather than Clang adding them >>> as metadata? >> >> >> Yep. >> >>> ValueTracking.cpp already has some iffy target knowledge (someone sneaked >>> a direct Intrinsic::x86_sse42_crc32_64_64 check in there), but extending it >>> to other intrinsics in other targets seems like too much... >> >> >> That's not iffy. That's exactly how it...

...d prefer LLVM passes >> >>> to >> >>> know about the range of these PTX intrinsics, rather than Clang adding >> >>> them >> >>> as metadata? >> >> >> >> >> >> Yep. >> >> >> >>> ValueTracking.cpp already has some iffy target knowledge (someone >> >>> sneaked >> >>> a direct Intrinsic::x86_sse42_crc32_64_64 check in there), but >> >>> extending it >> >>> to other intrinsics in other targets seems like too much... >> &g...

...at 6:26 AM Bozhenov, Nikolai <nikolai.bozhenov at intel.com> wrote: > > Hello, > > As I can see, recently LLVM switched to using assume bundles to encode alignment information: https://reviews.llvm.org/rG78de7297abe2e8fa782682168989c70e3cb34a5c > > However, it seems that the ValueTracking cannot understand the new format. As an example, consider compilation of the following reproducer with clang-11 (old assume format) and clang-trunk (assume bundles): > > #include <stdint.h> > int foo(int *p) { > __builtin_assume_aligned(p, 32); > return ((intptr_...

Is there any particular reason that CaptureTracking.h isn't folded into ValueTracking.h? I was looking for the method to check whether a pointer would be captured and looked in ValueTracking.h, unaware that a new header had been created for that one method. Would anyone mind if I folded CaptureTracking into ValueTracking? Nick -------------- next part -------------- An HTML att...

Hello, This simple patch fixes PR9112: Index: lib/Analysis/ValueTracking.cpp =================================================================== --- lib/Analysis/ValueTracking.cpp (revision 125281) +++ lib/Analysis/ValueTracking.cpp (working copy) @@ -593,6 +593,8 @@ // Otherwise take the unions of the known bit sets of the operands, // taking conse...

...he optimizer relies on `llvm::isSafeToSpeculativelyExecute` to decide whether the division should be performed speculatively. Right now, this function returns true. One could do something similar to integer divisions and add a case there (this solved the issue for me): ``` diff --git a/lib/Analysis/ValueTracking.cpp b/lib/Analysis/ValueTracking.cpp index 1761dac..c61fefd 100644 --- a/lib/Analysis/ValueTracking.cpp +++ b/lib/Analysis/ValueTracking.cpp @@ -3352,6 +3352,21 @@ bool llvm::isSafeToSpeculativelyExecute(const Value *V, // The numerator *might* be MinSignedValue. return false; } +...

...mplementation below and try to add some more context. > DominatorTree and AssumptionCache do seem to be sufficient and already > present. computeOverflowForSignedAdd and computeOverflowForUnsignedAdd > look like the kind of places to extend with this analysis based on > that info. ValueTracking has access to DominatorTree and AssumptionCache information, and provides the named two functions. InstCombine calls them and sets nsw/nuw flags based on the result. > Two questions about that approach though: > > 1. Going back to the predsimplify problems: performance. Walking up...

Eric Christopher wrote: > How are you trying to call it? Do you have a DataLayout? In test/Analysis/ValueTracking/memory-dereferenceable.ll, just change byval to dereferenceable(8), and %dparam won't match (see lib/IR/Value.cpp:521 for the logic that is supposed to fire). How do I get it to pass? I tried introducing a target-triple and target-datalayout, but it didn't help.

...ypes.h.in part might be controversial. Also, >> there's probably a better place to put it, but I'm not sure where. > > I didn't apply this part. What problems does it cause to not have > this? Can we fix uses of max and min? I get these errors in lib: .../lib/Analysis/ValueTracking.cpp:162: error: no matching function for call to 'min(unsigned int&, uint32_t&)' .../lib/Analysis/ValueTracking.cpp:163: error: no matching function for call to 'min(unsigned int&, uint32_t&)' .../lib/Analysis/ValueTracking.cpp:358: error: no matching function for...

...elyExecute` to decide whether the >> division should be performed speculatively. Right now, this function >> returns true. One could do something similar to integer divisions and >> add a case there (this solved the issue for me): >> ``` >> diff --git a/lib/Analysis/ValueTracking.cpp >> b/lib/Analysis/ValueTracking.cpp >> index 1761dac..c61fefd 100644 >> --- a/lib/Analysis/ValueTracking.cpp >> +++ b/lib/Analysis/ValueTracking.cpp >> @@ -3352,6 +3352,21 @@ bool llvm::isSafeToSpeculativelyExecute(const >> Value *V, >> // The...

On Jan 19, 2009, at 7:44 AM, Jay Foad wrote: > LLVM doesn't build out of the box on Cygwin, because Cygwin defines > uint32_t as "unsigned long" rather than "unsigned int". > > The attached patch allows me to build head of svn (rev 62510) on > Cygwin using GCC 4. Can this be committed please? Thanks applied:

I notice that GetUnderlyingObject has a few checks, but alloca isn't one of them. Then it fall backs to SimplifyInstruction which doesn't know about alloca so falls back to just trying to constant fold it. This seems a little silly since I assume alloca can't be constant folded. Should we just detect this early in GetUnderlyingObject and stop? ~Craig -------------- next part

LLVM doesn't build out of the box on Cygwin, because Cygwin defines uint32_t as "unsigned long" rather than "unsigned int". The attached patch allows me to build head of svn (rev 62510) on Cygwin using GCC 4. Can this be committed please? I realise that the DataTypes.h.in part might be controversial. Also, there's probably a better place to put it, but I'm not

Yep. Makes sense to me. There's nothing to simplify or constant-fold about an alloca. -Hal On 04/12/2017 04:23 PM, Craig Topper wrote: > Ping > > ~Craig > > On Fri, Apr 7, 2017 at 1:25 PM, Craig Topper <craig.topper at gmail.com > <mailto:craig.topper at gmail.com>> wrote: > > I notice that GetUnderlyingObject has a few checks, but alloca >