Separate more constant factors of parameters

  So far we separated constant factors from multiplications, however,
  only when they are at the outermost level of a parameter SCEV. Now,
  we also separate constant factors from the parameter SCEV if the
  outermost expression is a SCEVAddRecExpr. With the changes to the
  SCEVAffinator we can now improve the extractConstantFactor(...)
  function at will without worrying about any other code part. Thus,
  if needed we can implement a more comprehensive
  extractConstantFactor(...) function that will traverse the SCEV
  instead of looking only at the outermost level.

  Four test cases were affected. One did not change much and the other
  three were simplified.

llvm-svn: 260859

7 files changed, 59 insertions, 48 deletions

diff --git a/polly/include/polly/Support/SCEVValidator.h b/polly/include/polly/Support/SCEVValidator.h
index f58510a16cf..5d83d1397e0 100644
--- a/polly/include/polly/Support/SCEVValidator.h
+++ b/polly/include/polly/Support/SCEVValidator.h
@@ -19,6 +19,7 @@
 namespace llvm {
 class Region;
 class SCEV;
+class SCEVConstant;
 class ScalarEvolution;
 class Value;
 class Loop;
@@ -67,7 +68,7 @@ getParamsInAffineExpr(const llvm::Region *R, const llvm::SCEV *Expression,
 /// @param SE The ScalarEvolution analysis to create new SCEVs.
 ///
 /// @returns The constant factor in @p M and the rest of @p M.
-std::pair<const llvm::SCEV *, const llvm::SCEV *>
+std::pair<const llvm::SCEVConstant *, const llvm::SCEV *>
 extractConstantFactor(const llvm::SCEV *M, llvm::ScalarEvolution &SE);
 }
 
diff --git a/polly/lib/Support/SCEVAffinator.cpp b/polly/lib/Support/SCEVAffinator.cpp
index 2fee1e917a4..e94df79d19d 100644
--- a/polly/lib/Support/SCEVAffinator.cpp
+++ b/polly/lib/Support/SCEVAffinator.cpp
@@ -158,6 +158,10 @@ __isl_give isl_pw_aff *SCEVAffinator::visit(const SCEV *Expr) {
   if (PWA)
     return isl_pw_aff_copy(PWA);
 
+  auto ConstantAndLeftOverPair = extractConstantFactor(Expr, *S->getSE());
+  auto *Factor = ConstantAndLeftOverPair.first;
+  Expr = ConstantAndLeftOverPair.second;
+
   // In case the scev is a valid parameter, we do not further analyze this
   // expression, but create a new parameter in the isl_pw_aff. This allows us
   // to treat subexpressions that we cannot translate into an piecewise affine
@@ -171,18 +175,17 @@ __isl_give isl_pw_aff *SCEVAffinator::visit(const SCEV *Expr) {
     Affine = isl_aff_add_coefficient_si(Affine, isl_dim_param, 0, 1);
 
     PWA = isl_pw_aff_alloc(Domain, Affine);
-    CachedExpressions[Key] = PWA;
-    return isl_pw_aff_copy(PWA);
+  } else {
+    PWA = SCEVVisitor<SCEVAffinator, isl_pw_aff *>::visit(Expr);
   }
 
-  PWA = SCEVVisitor<SCEVAffinator, isl_pw_aff *>::visit(Expr);
+  PWA = isl_pw_aff_mul(visitConstant(Factor), PWA);
 
   // For compile time reasons we need to simplify the PWA before we cache and
   // return it.
   PWA = isl_pw_aff_coalesce(PWA);
-
-  CachedExpressions[Key] = PWA;
-  return isl_pw_aff_copy(PWA);
+  CachedExpressions[Key] = isl_pw_aff_copy(PWA);
+  return PWA;
 }
 
 __isl_give isl_pw_aff *SCEVAffinator::visitConstant(const SCEVConstant *Expr) {
@@ -235,15 +238,7 @@ __isl_give isl_pw_aff *SCEVAffinator::visitAddExpr(const SCEVAddExpr *Expr) {
 }
 
 __isl_give isl_pw_aff *SCEVAffinator::visitMulExpr(const SCEVMulExpr *Expr) {
-  // Divide Expr into a constant part and the rest. Then visit both and multiply
-  // the result to obtain the representation for Expr. While the second part of
-  // ConstantAndLeftOverPair might still be a SCEVMulExpr we will not get to
-  // this point again. The reason is that if it is a multiplication it consists
-  // only of parameters and we will stop in the visit(const SCEV *) function and
-  // return the isl_pw_aff for that parameter.
-  auto ConstantAndLeftOverPair = extractConstantFactor(Expr, *S->getSE());
-  return isl_pw_aff_mul(visit(ConstantAndLeftOverPair.first),
-                        visit(ConstantAndLeftOverPair.second));
+  llvm_unreachable("SCEVMulExpr should not be reached");
 }
 
 __isl_give isl_pw_aff *SCEVAffinator::visitUDivExpr(const SCEVUDivExpr *Expr) {
diff --git a/polly/lib/Support/SCEVValidator.cpp b/polly/lib/Support/SCEVValidator.cpp
index e22d3df162e..c32b2eef881 100644
--- a/polly/lib/Support/SCEVValidator.cpp
+++ b/polly/lib/Support/SCEVValidator.cpp
@@ -640,19 +640,35 @@ std::vector<const SCEV *> getParamsInAffineExpr(const Region *R,
   return Result.getParameters();
 }
 
-std::pair<const SCEV *, const SCEV *>
+std::pair<const SCEVConstant *, const SCEV *>
 extractConstantFactor(const SCEV *S, ScalarEvolution &SE) {
 
-  const SCEV *LeftOver = SE.getConstant(S->getType(), 1);
-  const SCEV *ConstPart = SE.getConstant(S->getType(), 1);
+  auto *LeftOver = SE.getConstant(S->getType(), 1);
+  auto *ConstPart = cast<SCEVConstant>(SE.getConstant(S->getType(), 1));
+
+  if (auto *Constant = dyn_cast<SCEVConstant>(S))
+    return std::make_pair(Constant, LeftOver);
+
+  auto *AddRec = dyn_cast<SCEVAddRecExpr>(S);
+  if (AddRec) {
+    auto *StartExpr = AddRec->getStart();
+    if (StartExpr->isZero()) {
+      auto StepPair = extractConstantFactor(AddRec->getStepRecurrence(SE), SE);
+      auto *LeftOverAddRec =
+          SE.getAddRecExpr(StartExpr, StepPair.second, AddRec->getLoop(),
+                           AddRec->getNoWrapFlags());
+      return std::make_pair(StepPair.first, LeftOverAddRec);
+    }
+    return std::make_pair(ConstPart, S);
+  }
 
-  const SCEVMulExpr *M = dyn_cast<SCEVMulExpr>(S);
-  if (!M)
+  auto *Mul = dyn_cast<SCEVMulExpr>(S);
+  if (!Mul)
     return std::make_pair(ConstPart, S);
 
-  for (const SCEV *Op : M->operands())
+  for (auto *Op : Mul->operands())
     if (isa<SCEVConstant>(Op))
-      ConstPart = SE.getMulExpr(ConstPart, Op);
+      ConstPart = cast<SCEVConstant>(SE.getMulExpr(ConstPart, Op));
     else
       LeftOver = SE.getMulExpr(LeftOver, Op);
 
diff --git a/polly/test/Isl/CodeGen/OpenMP/reference-preceeding-loop.ll b/polly/test/Isl/CodeGen/OpenMP/reference-preceeding-loop.ll
index e96a925728f..939b115da08 100644
--- a/polly/test/Isl/CodeGen/OpenMP/reference-preceeding-loop.ll
+++ b/polly/test/Isl/CodeGen/OpenMP/reference-preceeding-loop.ll
@@ -8,9 +8,9 @@
 ; AST:         {
 ; AST-NEXT:    #pragma simd
 ; AST-NEXT:    #pragma omp parallel for
-; AST-NEXT:    for (int c0 = 0; c0 < p_0 + symbol; c0 += 1)
+; AST-NEXT:    for (int c0 = 0; c0 < -p_0 + symbol; c0 += 1)
 ; AST-NEXT:      Stmt_while_body(c0);
-; AST-NEXT:    if (p_0 + symbol <= 0)
+; AST-NEXT:    if (p_0 >= symbol)
 ; AST-NEXT:      Stmt_while_body(0);
 ; AST-NEXT:    }
 
diff --git a/polly/test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_2.ll b/polly/test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_2.ll
index 491d38e409a..c10b2e603ba 100644
--- a/polly/test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_2.ll
+++ b/polly/test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_2.ll
@@ -21,15 +21,14 @@
 ; INNERMOST-NEXT: Invariant Accesses: {
 ; INNERMOST-NEXT: }
 ; INNERMOST-NEXT: Context:
-; INNERMOST-NEXT: [p_0, p_1, p_2, p_3] -> {  : 0 <= p_0 <= 1048576 and 0 <= p_1 <= 4096 and 0 <= p_2 <= 4096 and 0 <= p_3 <= 4194304 }
+; INNERMOST-NEXT: [p_0, p_1, p_2] -> {  : 0 <= p_0 <= 1048576 and 0 <= p_1 <= 1024 and 0 <= p_2 <= 1024 }
 ; INNERMOST-NEXT: Assumed Context:
-; INNERMOST-NEXT: [p_0, p_1, p_2, p_3] -> {  :  }
+; INNERMOST-NEXT: [p_0, p_1, p_2] -> {  :  }
 ; INNERMOST-NEXT: Boundary Context:
-; INNERMOST-NEXT: [p_0, p_1, p_2, p_3] -> {  :  }
+; INNERMOST-NEXT: [p_0, p_1, p_2] -> {  :  }
 ; INNERMOST-NEXT: p0: {0,+,{0,+,1}<nuw><nsw><%bb11>}<nuw><nsw><%bb13>
-; INNERMOST-NEXT: p1: {0,+,4}<nuw><nsw><%bb11>
-; INNERMOST-NEXT: p2: {0,+,4}<nuw><nsw><%bb13>
-; INNERMOST-NEXT: p3: {0,+,{0,+,4}<nuw><nsw><%bb11>}<%bb13>
+; INNERMOST-NEXT: p1: {0,+,1}<nuw><nsw><%bb11>
+; INNERMOST-NEXT: p2: {0,+,1}<nuw><nsw><%bb13>
 ; INNERMOST-NEXT: Arrays {
 ; INNERMOST-NEXT:     i32 MemRef_A[*]; // Element size 4
 ; INNERMOST-NEXT:     i64 MemRef_indvars_iv_next6; // Element size 8
@@ -45,26 +44,26 @@
 ; INNERMOST-NEXT: Statements {
 ; INNERMOST-NEXT:     Stmt_bb16
 ; INNERMOST-NEXT:         Domain :=
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb16[i0] : 0 <= i0 <= 1023 - p_0 };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] : 0 <= i0 <= 1023 - p_0 };
 ; INNERMOST-NEXT:         Schedule :=
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb16[i0] -> [0, i0] };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> [0, i0] };
 ; INNERMOST-NEXT:         ReadAccess :=    [Reduction Type: NONE] [Scalar: 0]
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb16[i0] -> MemRef_A[o0] : 4o0 = p_1 };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> MemRef_A[p_1] };
 ; INNERMOST-NEXT:         ReadAccess :=    [Reduction Type: NONE] [Scalar: 0]
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb16[i0] -> MemRef_A[o0] : 4o0 = p_2 };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> MemRef_A[p_2] };
 ; INNERMOST-NEXT:         ReadAccess :=    [Reduction Type: +] [Scalar: 0]
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb16[i0] -> MemRef_A[o0] : 4o0 = p_3 + 4i0 };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> MemRef_A[p_0 + i0] };
 ; INNERMOST-NEXT:         MustWriteAccess :=    [Reduction Type: +] [Scalar: 0]
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb16[i0] -> MemRef_A[o0] : 4o0 = p_3 + 4i0 };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> MemRef_A[p_0 + i0] };
 ; INNERMOST-NEXT:     Stmt_bb26
 ; INNERMOST-NEXT:         Domain :=
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb26[] : p_0 <= 1024 };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb26[] : p_0 <= 1024 };
 ; INNERMOST-NEXT:         Schedule :=
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb26[] -> [1, 0] };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb26[] -> [1, 0] };
 ; INNERMOST-NEXT:         MustWriteAccess :=    [Reduction Type: NONE] [Scalar: 1]
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb26[] -> MemRef_indvars_iv_next6[] };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb26[] -> MemRef_indvars_iv_next6[] };
 ; INNERMOST-NEXT:         MustWriteAccess :=    [Reduction Type: NONE] [Scalar: 1]
-; INNERMOST-NEXT:             [p_0, p_1, p_2, p_3] -> { Stmt_bb26[] -> MemRef_indvars_iv_next4[] };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb26[] -> MemRef_indvars_iv_next4[] };
 ; INNERMOST-NEXT: }
 
 ; ALL:      Function: f
diff --git a/polly/test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_3.ll b/polly/test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_3.ll
index 75ecbefd666..c744e8c683b 100644
--- a/polly/test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_3.ll
+++ b/polly/test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_3.ll
@@ -20,14 +20,14 @@
 ; INNERMOST-NEXT: Invariant Accesses: {
 ; INNERMOST-NEXT: }
 ; INNERMOST-NEXT: Context:
-; INNERMOST-NEXT: [p_0, p_1, p_2] -> {  : 0 <= p_0 <= 2147483647 and 0 <= p_1 <= 4096 and 0 <= p_2 <= 4096 }
+; INNERMOST-NEXT: [p_0, p_1, p_2] -> {  : 0 <= p_0 <= 2147483647 and 0 <= p_1 <= 1024 and 0 <= p_2 <= 1024 }
 ; INNERMOST-NEXT: Assumed Context:
 ; INNERMOST-NEXT: [p_0, p_1, p_2] -> {  :  }
 ; INNERMOST-NEXT: Boundary Context:
 ; INNERMOST-NEXT: [p_0, p_1, p_2] -> {  :  }
 ; INNERMOST-NEXT: p0: {0,+,{0,+,1}<nuw><nsw><%bb11>}<nuw><nsw><%bb13>
-; INNERMOST-NEXT: p1: {0,+,4}<nuw><nsw><%bb11>
-; INNERMOST-NEXT: p2: {0,+,4}<nuw><nsw><%bb13>
+; INNERMOST-NEXT: p1: {0,+,1}<nuw><nsw><%bb11>
+; INNERMOST-NEXT: p2: {0,+,1}<nuw><nsw><%bb13>
 ; INNERMOST-NEXT: Arrays {
 ; INNERMOST-NEXT:     i32 MemRef_A[*]; // Element size 4
 ; INNERMOST-NEXT:     i64 MemRef_indvars_iv_next6; // Element size 8
@@ -47,9 +47,9 @@
 ; INNERMOST-NEXT:         Schedule :=
 ; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> [0, i0] };
 ; INNERMOST-NEXT:         ReadAccess :=    [Reduction Type: NONE] [Scalar: 0]
-; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> MemRef_A[o0] : 4o0 = p_1 };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> MemRef_A[p_1] };
 ; INNERMOST-NEXT:         ReadAccess :=    [Reduction Type: NONE] [Scalar: 0]
-; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> MemRef_A[o0] : 4o0 = p_2 };
+; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> MemRef_A[p_2] };
 ; INNERMOST-NEXT:         ReadAccess :=    [Reduction Type: +] [Scalar: 0]
 ; INNERMOST-NEXT:             [p_0, p_1, p_2] -> { Stmt_bb16[i0] -> MemRef_A[i0] };
 ; INNERMOST-NEXT:         MustWriteAccess :=    [Reduction Type: +] [Scalar: 0]
diff --git a/polly/test/ScopInfo/constant_start_integer.ll b/polly/test/ScopInfo/constant_start_integer.ll
index 260b77d0592..e306d1660dd 100644
--- a/polly/test/ScopInfo/constant_start_integer.ll
+++ b/polly/test/ScopInfo/constant_start_integer.ll
@@ -15,9 +15,9 @@ target datalayout = "e-p:64:64:64-i1:8:8-i8:8:8-i16:16:16-i32:32:32-i64:64:64-f3
 ; CHECK-NOT: p1
 
 ; CHECK: ReadAccess
-; CHECK:   [p_0] -> { Stmt_for_body3[i0] -> MemRef_input[o0] : 4o0 = 4 + p_0 + 4i0 };
+; CHECK:   [p_0] -> { Stmt_for_body3[i0] -> MemRef_input[1 + 64p_0 + i0] };
 ; CHECK: MustWriteAccess
-; CHECK:   [p_0] -> { Stmt_for_body3[i0] -> MemRef_input[o0] : 4o0 = p_0 + 4i0 };
+; CHECK:   [p_0] -> { Stmt_for_body3[i0] -> MemRef_input[64p_0 + i0] };
 
 define void @foo(float* nocapture %input) {
 entry: