[Polly] Isl AST reduction pragmas [V2]

[Tobias Grosser]

:
On Jun 20, 2014 6:27 PM, Johannes Doerfert <jdoerfert at codeaurora.org> wrote:
>
> I get your point now, thank you! 
>
> I was all the time thinking about the current scheduler and legal transformations, but if we remove reduction dependences such transformations become legal too... 
>
> Just an Idea I had on the way to work: 
>   Can we use the transitive hull (after removing backward dependences) and the reverse of that to check if reduction dependences are broken? 

Right. Just make sure to first reverse and then remove the negative dependences. Also it may make sense to always keep the full set of dependences and only remove dependences on demand. Like this the dependences won't be incorrect in case the schedule changes.

>   We would handle at least the two example (loop reversal and the modulo distribution) correct. 
>
> Best regards and thanks for being so patient with me :)

Same for you. I learn a lot about reductions these days.

Tobias 
>   Johannes 
>
> -- 
>
> Johannes Doerfert 
> Employee of Qualcomm Innovation Center, Inc. 
> Qualcomm Innovation Center, Inc. is a member of Code Aurora Forum, hosted by The Linux Foundation 
>
> -----Original Message----- 
> From: Tobias Grosser [mailto:tobias at grosser.es] 
> Sent: Friday, June 20, 2014 1:26 AM 
> To: Johannes Doerfert; 'Johannes Doerfert' 
> Cc: llvm-commits at cs.uiuc.edu; 'Sebastian Pop' 
> Subject: Re: [Polly] Isl AST reduction pragmas [V2] 
>
> On 19/06/2014 20:47, Johannes Doerfert wrote: 
> > On 19/06/2014 18:56, Johannes Doerfert wrote: 
> >> So if I have the dependences and schedule like shown below (assume they are e.g., WAW dependences). 
> >> Would a function like "astScheduleDimIsParallel" in IslAst tell me it is parallel or not (without my patch and reductions)? 
> >> 
> >> I was under the impression that it would (and should) say __not__ parallel. 
> > 
> >>> It would say 'parallel', as there are no loop carried dependences between the statements that are part of the loop. Or does the code suggest otherwise? 
> > There is a loop carried dependency. 
> > 
> > Looking at the C versions (before & after) we clearly see the assignments to the same memory location. 
> > Looking at the dependences we see a dependency between one iteration and every following one (even if transitive dependences are only implicitly represented). 
> > Looking at the "parallelism tests" (e.g., the one ion astScheduleDimIsParallel [btw. there are also others which perform the same/a similar function... we should not duplicate such important functions!]), 
> >    we see that the original dependences S[i]->S[i+1] are transformed according to the schedule and then we look if the dependency is carried by an outer loop or not carried by the current dimension. 
> > 
> > My point is that in all these cases I see the forward dependency from one iteration to the next of the same loop, and one between the loops. 
> > 
> > I simply do not understand where we get parallelism here, other than from the "excluded" (see above) reductions. Even with reductions the test (to my understanding) will yield the correct result (which is now "reduction parallel", not "not parallel" any more). 
>
> I attached a test case derived from this C file. 
>
> void foo(float *A, long n) { 
> for (long i = 0; i < 2*n; i++) 
>   A[0] += i; 
>
> for (long i = 0; i < 2*n; i++) 
>            A[i + 1] = 1; 
> } 
>
> $ polly-opt -polly-ast -analyze test.preopt.ll -disable-polly-legality -polly-ast-detect-parallel 
>
> if (1) 
>
>      { 
>        for (int c1 = 0; c1 < 2 * n; c1 += 1) 
>          Stmt_for_body(c1); 
>        #pragma simd 
>        #pragma omp parallel for 
>        for (int c1 = 0; c1 < 2 * n; c1 += 1) 
>          Stmt_for_body6(c1); 
>      } 
>
> else 
>      {  /* original code */ } 
>
>
> We can import the following schedule (see attached jscop file): 
>
>
> "[n] -> { Stmt_for_body[i0] -> scattering[0, i0, 0]: i0 % 2 = 0; 
>            Stmt_for_body[i0] -> scattering[2, i0, 0]: i0 % 2 = 1 }" 
> "[n] -> { Stmt_for_body6[i0] -> scattering[1, i0, 0] }" 
>
>
> $ polly-opt -polly-import-jscop -polly-ast -analyze test.preopt.ll -disable-polly-legality -polly-ast-detect-parallel 
>
>
>      for (int c0 = 0; c0 <= 2; c0 += 1) { 
>        if (c0 == 2) { 
>          #pragma simd 
>          #pragma omp parallel for 
>          for (int c1 = 1; c1 < 2 * n; c1 += 2) 
>            Stmt_for_body(c1); 
>        } else if (c0 == 1) { 
>          #pragma simd 
>          #pragma omp parallel for 
>          for (int c1 = 0; c1 < 2 * n; c1 += 1) 
>            Stmt_for_body6(c1); 
>        } else 
>          #pragma simd 
>          #pragma omp parallel for 
>          for (int c1 = 0; c1 < 2 * n - 1; c1 += 2) 
>            Stmt_for_body(c1); 
>      } 
>
> Cheers, 
> Tobias 
>