[PATCH] D11311: Replace hardcoded threshold with an option when analyze read attributes

Nick Lewycky nicholas at mxc.ca
Mon Jul 20 21:24:56 PDT 2015


nicholas added a comment.

Xuetian Weng wrote:

> wengxt added a comment.

> 

> > Question one, what does this have to do with ld.global.nc? It looks like

> 

> 

> 

> 

> >   you've got a load instruction (wrapped in an intrinsic) and it in turn

> 

> 

> 

> 

> >   has many uses. The fact it's ld.global.nc doesn't seem to be relevant?

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > In theory, if you replaced the ld.global.nc with a LoadInst, would you

> 

> 

> 

> 

> >   have the same problem? Or is there some other optimization that occurs

> 

> 

> 

> 

> >   on those uses of a LoadInst that doesn't occur on the ld.global.nc?

> 

> 

> In our case, we intend to transform LoadInst into ld.global.nc during Instruction

>  selection, not the other direction. It's only safe to use ld.global.nc for LoadInst

>  if the pointer being accessed is readonly (and noalias, which is irreverent to this

>  patch).


I realize that, but my question stands. Is there an optimization that 
should be happening to these ld.global.nc before we get here?

> > Question two, can you think of a way to design determinePointerReadAttrs

> 

> 

> 

> 

> >   so that it doesn't need a use count but doesn't have a bad worst case

> 

> 

> 

> 

> >   performance problem? For that matter, why isn't this code already linear

> 

> 

> 

> 

> >   in the number of instructions?

> 

> 

> Yeah, I also think this is linear in the number of instructions, and I don't think

>  there can be multiple call on this function that make it quadratic or something.

>  The complexity should be O(num_of_arguments * num_of_instructions)

>  Maybe simply remove this threshold is good enough. Is there any other reason

>  to introduce this threshold?


Nope, I can't think of any reason. I think this threshold is safe to remove!

Nick

> > > [1] https://github.com/vetter/shoc

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > http://reviews.llvm.org/D11311

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > Files:

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > >   lib/Transforms/IPO/FunctionAttrs.cpp

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > >   test/Transforms/FunctionAttrs/readattrs-threshold.ll

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > Index: test/Transforms/FunctionAttrs/readattrs-threshold.ll

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > >   ===================================================================

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > - /dev/null +++ test/Transforms/FunctionAttrs/readattrs-threshold.ll @@ -0,0 +1,13 @@ +; RUN: opt<   %s -functionattrs -S | FileCheck %s --check-prefix=CHECK +; RUN: opt<   %s -functionattrs -readattrs-analysis-uses-threshold=1 -S | FileCheck %s --check-prefix=CHECK-NO-READATTR + +; %p has two uses +; CHECK: define i32 @test(i32* nocapture readonly %p) +; CHECK-NO-READATTR: define i32 @test(i32* nocapture %p) +define i32 @test(i32* %p) { +  %1 = load i32, i32* %p +  %2 = getelementptr i32, i32* %p, i32 1 +  %3 = load i32, i32* %2 +  %4 = add i32 %1, %3 +  ret i32 %4 +} Index: lib/Transforms/IPO/FunctionAttrs.cpp ===================================================================

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > - lib/Transforms/IPO/FunctionAttrs.cpp +++ lib/Transforms/IPO/FunctionAttrs.cpp @@ -44,6 +44,12 @@ STATISTIC(NumNoAlias, "Number of function returns marked noalias"); STATISTIC(NumAnnotated, "Number of attributes added to library functions");

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > >   +static cl::opt<int>   ReadAttrsAnalysisUsesThreshold( +    "readattrs-analysis-uses-threshold", cl::Hidden, cl::init(20), +    cl::ZeroOrMore, +    cl::desc("Control the maximum number of uses to analyze for " +             "readonly/readnone attribute (default = 20)")); + namespace { struct FunctionAttrs : public CallGraphSCCPass { static char ID; // Pass identification, replacement for typeid @@ -425,7 +431,7 @@ // We don't need to track IsWritten. If A is written to, return immediately.

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > >   for (Use&U : A->uses()) {

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > - if (Count++>= 20) +    if (Count++>= ReadAttrsAnalysisUsesThreshold) return Attribute::None;

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > >   Visited.insert(&U);

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > >    _______________________________________________

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > 

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > > llvm-commits mailing list

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > >   llvm-commits at cs.uiuc.edu

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> 

> 

> > >   http://lists.cs.uiuc.edu/mailman/listinfo/llvm-commits

> 

> > 

> 

> 

> 

> 

> > 

> 

> 

> http://reviews.llvm.org/D11311



http://reviews.llvm.org/D11311







More information about the llvm-commits mailing list