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

diff --git a/polly/include/polly/Support/SCEVValidator.h b/polly/include/polly/Support/SCEVValidator.h
index f58510a..5d83d13 100644 (file)
--- 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 2fee1e9..e94df79 100644 (file)
--- 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 e22d3df..c32b2ee 100644 (file)
--- 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 e96a925..939b115 100644 (file)
--- 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 491d38e..c10b2e6 100644 (file)
--- 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 75ecbef..c744e8c 100644 (file)
--- 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 260b77d..e306d16 100644 (file)
--- 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: