Index: include/polly/ScopInfo.h =================================================================== --- include/polly/ScopInfo.h +++ include/polly/ScopInfo.h @@ -1306,16 +1306,12 @@ /// this scop and that need to be code generated as a run-time test. isl_set *AssumedContext; - /// @brief The boundary assumptions under which this scop was built. + /// @brief The restrictions under which this scop was built. /// - /// The boundary context is similar to the assumed context as it contains + /// The invalid context is similar to the assumed context as it contains /// constraints over the parameters we assume to be true. However, the - /// boundary context is less useful for dependence analysis and - /// simplification purposes as it contains only constraints that affect the - /// boundaries of the parameter ranges. As these constraints can become quite - /// complex, the boundary context and the assumed context are separated as a - /// meassure to save compile time. - isl_set *BoundaryContext; + /// invalid context ... TODO + isl_set *InvalidContext; /// @brief The schedule of the SCoP /// @@ -1522,8 +1518,8 @@ /// @brief Build the Context of the Scop. void buildContext(); - /// @brief Build the BoundaryContext based on the wrapping of expressions. - void buildBoundaryContext(); + /// @brief Add the restrictions based on the wrapping of expressions. + void addWrappingContext(); /// @brief Add user provided parameter constraints to context (source code). void addUserAssumptions(AssumptionCache &AC, DominatorTree &DT, LoopInfo &LI); @@ -1534,7 +1530,7 @@ /// @brief Add the bounds of the parameters to the context. void addParameterBounds(); - /// @brief Simplify the assumed and boundary context. + /// @brief Simplify the assumed and invalid context. void simplifyContexts(); /// @brief Get the representing SCEV for @p S if applicable, otherwise @p S. @@ -1762,14 +1758,6 @@ /// @return The assumed context of this Scop. __isl_give isl_set *getAssumedContext() const; - /// @brief Get the runtime check context for this Scop. - /// - /// The runtime check context contains all constraints that have to - /// hold at runtime for the optimized version to be executed. - /// - /// @return The runtime check context of this Scop. - __isl_give isl_set *getRuntimeCheckContext() const; - /// @brief Return true if the optimized SCoP can be executed. /// /// In addition to the runtime check context this will also utilize the domain @@ -1780,14 +1768,18 @@ /// @brief Track and report an assumption. /// - /// Use 'clang -Rpass-analysis=polly-scops' or 'opt -pass-remarks=polly-scops' - /// to output the assumptions. + /// Use 'clang -Rpass-analysis=polly-scops' or 'opt + /// -pass-remarks-analysis=polly-scops' to output the assumptions. /// - /// @param Kind The assumption kind describing the underlying cause. - /// @param Set The relations between parameters that are assumed to hold. - /// @param Loc The location in the source that caused this assumption. - void trackAssumption(AssumptionKind Kind, __isl_keep isl_set *Set, - DebugLoc Loc); + /// @param Kind The assumption kind describing the underlying cause. + /// @param Set The relations between parameters that are assumed to hold. + /// @param Loc The location in the source that caused this assumption. + /// @param IsPos Flag to indicate if the assumptions in @p Set are positive + /// (needed) or negative (invalid). + /// + /// @returns True if the assumption is not trivial. + bool trackAssumption(AssumptionKind Kind, __isl_keep isl_set *Set, + DebugLoc Loc, bool IsPos); /// @brief Add assumptions to assumed context. /// @@ -1803,8 +1795,10 @@ /// @param Kind The assumption kind describing the underlying cause. /// @param Set The relations between parameters that are assumed to hold. /// @param Loc The location in the source that caused this assumption. - void addAssumption(AssumptionKind Kind, __isl_take isl_set *Set, - DebugLoc Loc); + /// @param IsPos Flag to indicate if the assumptions in @p Set are positive + /// (needed) or negative (invalid). + void addAssumption(AssumptionKind Kind, __isl_take isl_set *Set, DebugLoc Loc, + bool IsPos); /// @brief Mark the scop as invalid. /// @@ -1817,10 +1811,10 @@ /// @param Loc The location in the source that triggered . void invalidate(AssumptionKind Kind, DebugLoc Loc); - /// @brief Get the boundary context for this Scop. + /// @brief Get the invalid context for this Scop. /// - /// @return The boundary context of this Scop. - __isl_give isl_set *getBoundaryContext() const; + /// @return The invalid context of this Scop. + __isl_give isl_set *getInvalidContext() const; /// @brief Build the alias checks for this SCoP. void buildAliasChecks(AliasAnalysis &AA); @@ -1841,8 +1835,8 @@ /// @brief Get an isl string representing the assumed context. std::string getAssumedContextStr() const; - /// @brief Get an isl string representing the boundary context. - std::string getBoundaryContextStr() const; + /// @brief Get an isl string representing the invalid context. + std::string getInvalidContextStr() const; /// @brief Return the stmt for the given @p BB or nullptr if none. ScopStmt *getStmtForBasicBlock(BasicBlock *BB) const; Index: lib/Analysis/ScopInfo.cpp =================================================================== --- lib/Analysis/ScopInfo.cpp +++ lib/Analysis/ScopInfo.cpp @@ -589,9 +589,9 @@ // bail out more often than strictly necessary. Outside = isl_set_remove_divs(Outside); Outside = isl_set_complement(Outside); - Statement->getParent()->addAssumption( - INBOUNDS, Outside, - getAccessInstruction() ? getAccessInstruction()->getDebugLoc() : nullptr); + auto &Loc = getAccessInstruction() ? getAccessInstruction()->getDebugLoc() + : DebugLoc(); + Statement->getParent()->addAssumption(INBOUNDS, Outside, Loc, true); isl_space_free(Space); } @@ -1278,6 +1278,7 @@ assert(IndexOffset <= 1 && "Unexpected large index offset"); + auto *NotExecuted = isl_set_complement(isl_set_params(getDomain())); for (size_t i = 0; i < Sizes.size(); i++) { auto Expr = Subscripts[i + IndexOffset]; auto Size = Sizes[i]; @@ -1305,17 +1306,17 @@ OutOfBound = isl_set_intersect(getDomain(), OutOfBound); OutOfBound = isl_set_params(OutOfBound); isl_set *InBound = isl_set_complement(OutOfBound); - isl_set *Executed = isl_set_params(getDomain()); // A => B == !A or B isl_set *InBoundIfExecuted = - isl_set_union(isl_set_complement(Executed), InBound); + isl_set_union(isl_set_copy(NotExecuted), InBound); InBoundIfExecuted = isl_set_coalesce(InBoundIfExecuted); - Parent.addAssumption(INBOUNDS, InBoundIfExecuted, GEP->getDebugLoc()); + Parent.addAssumption(INBOUNDS, InBoundIfExecuted, GEP->getDebugLoc(), true); } isl_local_space_free(LSpace); + isl_set_free(NotExecuted); } void ScopStmt::deriveAssumptions(BasicBlock *Block, ScopDetection &SD) { @@ -1715,37 +1716,12 @@ return isl_set_intersect_params(C, DomainContext); } -void Scop::buildBoundaryContext() { - if (IgnoreIntegerWrapping) { - BoundaryContext = isl_set_universe(getParamSpace()); +void Scop::addWrappingContext() { + if (IgnoreIntegerWrapping) return; - } - - BoundaryContext = Affinator.getWrappingContext(); - - // The isl_set_complement operation used to create the boundary context - // can possibly become very expensive. We bound the compile time of - // this operation by setting a compute out. - // - // TODO: We can probably get around using isl_set_complement and directly - // AST generate BoundaryContext. - long MaxOpsOld = isl_ctx_get_max_operations(getIslCtx()); - isl_ctx_reset_operations(getIslCtx()); - isl_ctx_set_max_operations(getIslCtx(), 300000); - isl_options_set_on_error(getIslCtx(), ISL_ON_ERROR_CONTINUE); - - BoundaryContext = isl_set_complement(BoundaryContext); - - if (isl_ctx_last_error(getIslCtx()) == isl_error_quota) { - isl_set_free(BoundaryContext); - BoundaryContext = isl_set_empty(getParamSpace()); - } - isl_options_set_on_error(getIslCtx(), ISL_ON_ERROR_ABORT); - isl_ctx_reset_operations(getIslCtx()); - isl_ctx_set_max_operations(getIslCtx(), MaxOpsOld); - BoundaryContext = isl_set_gist_params(BoundaryContext, getContext()); - trackAssumption(WRAPPING, BoundaryContext, DebugLoc()); + auto *WrappingContext = Affinator.getWrappingContext(); + addAssumption(WRAPPING, WrappingContext, DebugLoc(), false); } void Scop::addUserAssumptions(AssumptionCache &AC, DominatorTree &DT, @@ -1854,6 +1830,7 @@ void Scop::buildContext() { isl_space *Space = isl_space_params_alloc(IslCtx, 0); Context = isl_set_universe(isl_space_copy(Space)); + InvalidContext = isl_set_empty(isl_space_copy(Space)); AssumedContext = isl_set_universe(Space); } @@ -1933,7 +1910,7 @@ // otherwise we would access out of bound data. Now, knowing that code is // only executed for the case m >= 0, it is sufficient to assume p >= 0. AssumedContext = simplifyAssumptionContext(AssumedContext, *this); - BoundaryContext = simplifyAssumptionContext(BoundaryContext, *this); + InvalidContext = simplifyAssumptionContext(InvalidContext, *this); } /// @brief Add the minimal/maximal access in @p Set to @p User. @@ -2403,8 +2380,8 @@ if (containsErrorBlock(RN, getRegion(), LI, DT)) { IsOptimized = true; isl_set *DomPar = isl_set_params(isl_set_copy(Domain)); - addAssumption(ERRORBLOCK, isl_set_complement(DomPar), - BB->getTerminator()->getDebugLoc()); + addAssumption(ERRORBLOCK, DomPar, BB->getTerminator()->getDebugLoc(), + false); } } } @@ -2504,9 +2481,8 @@ } isl_set *UnboundedCtx = isl_set_params(Parts.first); - isl_set *BoundedCtx = isl_set_complement(UnboundedCtx); - addAssumption(INFINITELOOP, BoundedCtx, - HeaderBB->getTerminator()->getDebugLoc()); + addAssumption(INFINITELOOP, UnboundedCtx, + HeaderBB->getTerminator()->getDebugLoc(), false); } void Scop::buildAliasChecks(AliasAnalysis &AA) { @@ -2740,7 +2716,7 @@ : SE(&ScalarEvolution), R(R), IsOptimized(false), HasSingleExitEdge(R.getExitingBlock()), HasErrorBlock(false), MaxLoopDepth(MaxLoopDepth), IslCtx(Context), Context(nullptr), - Affinator(this), AssumedContext(nullptr), BoundaryContext(nullptr), + Affinator(this), AssumedContext(nullptr), InvalidContext(nullptr), Schedule(nullptr) { buildContext(); } @@ -2764,14 +2740,14 @@ buildSchedule(SD, LI); - if (isl_set_is_empty(AssumedContext)) + if (!hasFeasibleRuntimeContext()) return; updateAccessDimensionality(); realignParams(); addParameterBounds(); addUserContext(); - buildBoundaryContext(); + addWrappingContext(); simplifyContexts(); buildAliasChecks(AA); @@ -2782,7 +2758,7 @@ Scop::~Scop() { isl_set_free(Context); isl_set_free(AssumedContext); - isl_set_free(BoundaryContext); + isl_set_free(InvalidContext); isl_schedule_free(Schedule); for (auto It : DomainMap) @@ -3077,10 +3053,11 @@ std::string Scop::getContextStr() const { return stringFromIslObj(Context); } std::string Scop::getAssumedContextStr() const { + assert(AssumedContext && "Assumed context not yet built"); return stringFromIslObj(AssumedContext); } -std::string Scop::getBoundaryContextStr() const { - return stringFromIslObj(BoundaryContext); +std::string Scop::getInvalidContextStr() const { + return stringFromIslObj(InvalidContext); } std::string Scop::getNameStr() const { @@ -3106,22 +3083,30 @@ } __isl_give isl_set *Scop::getAssumedContext() const { + assert(AssumedContext && "Assumed context not yet built"); return isl_set_copy(AssumedContext); } -__isl_give isl_set *Scop::getRuntimeCheckContext() const { - isl_set *RuntimeCheckContext = getAssumedContext(); - RuntimeCheckContext = - isl_set_intersect(RuntimeCheckContext, getBoundaryContext()); - RuntimeCheckContext = simplifyAssumptionContext(RuntimeCheckContext, *this); - return RuntimeCheckContext; -} - bool Scop::hasFeasibleRuntimeContext() const { - isl_set *RuntimeCheckContext = getRuntimeCheckContext(); - RuntimeCheckContext = addNonEmptyDomainConstraints(RuntimeCheckContext); - bool IsFeasible = !isl_set_is_empty(RuntimeCheckContext); - isl_set_free(RuntimeCheckContext); + isl_set *PositiveContext = getAssumedContext(); + if (!PositiveContext) + return false; + + PositiveContext = addNonEmptyDomainConstraints(PositiveContext); + bool IsFeasible = !isl_set_is_empty(PositiveContext); + isl_set_free(PositiveContext); + if (!IsFeasible) + return false; + + isl_set *NegativeContext = getInvalidContext(); + if (!NegativeContext) + return false; + + isl_set *DomainContext = isl_union_set_params(getDomains()); + IsFeasible = !isl_set_is_subset(DomainContext, NegativeContext); + isl_set_free(NegativeContext); + isl_set_free(DomainContext); + return IsFeasible; } @@ -3149,67 +3134,62 @@ llvm_unreachable("Unknown AssumptionKind!"); } -void Scop::trackAssumption(AssumptionKind Kind, __isl_keep isl_set *Set, - DebugLoc Loc) { - if (isl_set_is_subset(Context, Set)) - return; +bool Scop::trackAssumption(AssumptionKind Kind, __isl_keep isl_set *Set, + DebugLoc Loc, bool IsPos) { + if (IsPos) { + if (isl_set_is_subset(Context, Set)) + return false; - if (isl_set_is_subset(AssumedContext, Set)) - return; + if (isl_set_is_subset(AssumedContext, Set)) + return false; + } else { + if (isl_set_is_disjoint(Set, Context)) + return false; + + if (isl_set_is_subset(Set, InvalidContext)) + return false; + } auto &F = *getRegion().getEntry()->getParent(); - std::string Msg = toString(Kind) + " assumption:\t" + stringFromIslObj(Set); + auto Suffix = IsPos ? " assumption:\t" : " restriction:\t"; + std::string Msg = toString(Kind) + Suffix + stringFromIslObj(Set); emitOptimizationRemarkAnalysis(F.getContext(), DEBUG_TYPE, F, Loc, Msg); + return true; } void Scop::addAssumption(AssumptionKind Kind, __isl_take isl_set *Set, - DebugLoc Loc) { - trackAssumption(Kind, Set, Loc); - AssumedContext = isl_set_intersect(AssumedContext, Set); - - int NSets = isl_set_n_basic_set(AssumedContext); - if (NSets >= MaxDisjunctsAssumed) { - isl_space *Space = isl_set_get_space(AssumedContext); - isl_set_free(AssumedContext); - AssumedContext = isl_set_empty(Space); + DebugLoc Loc, bool IsPos) { + if (!trackAssumption(Kind, Set, Loc, IsPos)) { + isl_set_free(Set); + return; } - AssumedContext = isl_set_coalesce(AssumedContext); + if (IsPos) { + AssumedContext = isl_set_intersect(AssumedContext, Set); + AssumedContext = isl_set_coalesce(AssumedContext); + } else { + InvalidContext = isl_set_union(InvalidContext, Set); + InvalidContext = isl_set_coalesce(InvalidContext); + } } void Scop::invalidate(AssumptionKind Kind, DebugLoc Loc) { - addAssumption(Kind, isl_set_empty(getParamSpace()), Loc); + addAssumption(Kind, isl_set_empty(getParamSpace()), Loc, true); } -__isl_give isl_set *Scop::getBoundaryContext() const { - return isl_set_copy(BoundaryContext); +__isl_give isl_set *Scop::getInvalidContext() const { + return isl_set_copy(InvalidContext); } void Scop::printContext(raw_ostream &OS) const { OS << "Context:\n"; - - if (!Context) { - OS.indent(4) << "n/a\n\n"; - return; - } - - OS.indent(4) << getContextStr() << "\n"; + OS.indent(4) << Context << "\n"; OS.indent(4) << "Assumed Context:\n"; - if (!AssumedContext) { - OS.indent(4) << "n/a\n\n"; - return; - } - - OS.indent(4) << getAssumedContextStr() << "\n"; - - OS.indent(4) << "Boundary Context:\n"; - if (!BoundaryContext) { - OS.indent(4) << "n/a\n\n"; - return; - } + OS.indent(4) << AssumedContext << "\n"; - OS.indent(4) << getBoundaryContextStr() << "\n"; + OS.indent(4) << "Invalid Context:\n"; + OS.indent(4) << InvalidContext << "\n"; for (const SCEV *Parameter : Parameters) { int Dim = ParameterIds.find(Parameter)->second; Index: lib/CodeGen/IslAst.cpp =================================================================== --- lib/CodeGen/IslAst.cpp +++ lib/CodeGen/IslAst.cpp @@ -332,8 +332,11 @@ // The conditions that need to be checked at run-time for this scop are // available as an isl_set in the runtime check context from which we can // directly derive a run-time condition. - RunCondition = - isl_ast_build_expr_from_set(Build, S->getRuntimeCheckContext()); + auto *OneV = isl_val_one(isl_ast_build_get_ctx(Build)); + auto *PosCond = isl_ast_build_expr_from_set(Build, S->getAssumedContext()); + auto *NegCond = isl_ast_build_expr_from_set(Build, S->getInvalidContext()); + auto *NotNegCond = isl_ast_expr_sub(isl_ast_expr_from_val(OneV), NegCond); + RunCondition = isl_ast_expr_and(PosCond, NotNegCond); // Create the alias checks from the minimal/maximal accesses in each alias // group which consists of read only and non read only (read write) accesses. Index: test/Isl/Ast/OpenMP/nested_loop_both_parallel_parametric.ll =================================================================== --- test/Isl/Ast/OpenMP/nested_loop_both_parallel_parametric.ll +++ test/Isl/Ast/OpenMP/nested_loop_both_parallel_parametric.ll @@ -40,7 +40,7 @@ ret void } -; CHECK: if (n <= 1024) +; CHECK: if (n <= 1024 && 1 - 0) ; CHECK: #pragma omp parallel for ; CHECK: for (int c0 = 0; c0 < n; c0 += 1) ; CHECK: #pragma simd Index: test/Isl/Ast/alias_simple_1.ll =================================================================== --- test/Isl/Ast/alias_simple_1.ll +++ test/Isl/Ast/alias_simple_1.ll @@ -12,11 +12,11 @@ ; A[i] = B[i]; ; } ; -; NOAA: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) -; BASI: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) -; TBAA: if (1) -; SCEV: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) -; GLOB: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; NOAA: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; BASI: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; TBAA: if (1 && 1 - 0) +; SCEV: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; GLOB: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) ; target datalayout = "e-m:e-i64:64-f80:128-n8:16:32:64-S128" Index: test/Isl/Ast/alias_simple_2.ll =================================================================== --- test/Isl/Ast/alias_simple_2.ll +++ test/Isl/Ast/alias_simple_2.ll @@ -12,11 +12,11 @@ ; A[i] = B[i]; ; } ; -; NOAA: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) -; BASI: if (1) -; TBAA: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) -; SCEV: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) -; GLOB: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; NOAA: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; BASI: if (1 && 1 - 0) +; TBAA: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; SCEV: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; GLOB: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) ; target datalayout = "e-m:e-i64:64-f80:128-n8:16:32:64-S128" Index: test/Isl/Ast/alias_simple_3.ll =================================================================== --- test/Isl/Ast/alias_simple_3.ll +++ test/Isl/Ast/alias_simple_3.ll @@ -12,11 +12,11 @@ ; A[i] = B[i]; ; } ; -; NOAA: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) -; BASI: if (1) -; TBAA: if (1) -; SCEV: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) -; GLOB: if (1 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; NOAA: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; BASI: if (1 && 1 - 0) +; TBAA: if (1 && 1 - 0) +; SCEV: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) +; GLOB: if (1 && 1 - 0 && (&MemRef_B[N] <= &MemRef_A[0] || &MemRef_A[N] <= &MemRef_B[0])) ; target datalayout = "e-m:e-i64:64-f80:128-n8:16:32:64-S128" Index: test/Isl/Ast/aliasing_multiple_alias_groups.ll =================================================================== --- test/Isl/Ast/aliasing_multiple_alias_groups.ll +++ test/Isl/Ast/aliasing_multiple_alias_groups.ll @@ -8,7 +8,7 @@ ; } ; } ; -; NOAA: if (1 && ( +; NOAA: if (1 && 1 - 0 && ( ; NOAA-DAG: &MemRef_Int0[1024] <= &MemRef_Float0[0] || &MemRef_Float0[1024] <= &MemRef_Int0[0] ; NOAA-DAG: &MemRef_Int1[1024] <= &MemRef_Float0[0] || &MemRef_Float0[1024] <= &MemRef_Int1[0] ; NOAA-DAG: &MemRef_Float1[1024] <= &MemRef_Float0[0] || &MemRef_Float0[1024] <= &MemRef_Float1[0] @@ -16,7 +16,7 @@ ; NOAA-DAG: &MemRef_Float1[1024] <= &MemRef_Int0[0] || &MemRef_Int0[1024] <= &MemRef_Float1[0] ; NOAA: )) ; -; TBAA: if (1 && ( +; TBAA: if (1 && 1 - 0 && ( ; TBAA-DAG: &MemRef_Int1[1024] <= &MemRef_Int0[0] || &MemRef_Int0[1024] <= &MemRef_Int1[0] ; TBAA-DAG: &MemRef_Float1[1024] <= &MemRef_Float0[0] || &MemRef_Float0[1024] <= &MemRef_Float1[0] ; TBAA: )) Index: test/Isl/Ast/aliasing_parametric_simple_1.ll =================================================================== --- test/Isl/Ast/aliasing_parametric_simple_1.ll +++ test/Isl/Ast/aliasing_parametric_simple_1.ll @@ -5,7 +5,7 @@ ; A[i] = B[c]; ; } ; -; CHECK: if (1 && (&MemRef_B[c + 1] <= &MemRef_A[0] || &MemRef_A[1024] <= &MemRef_B[c])) +; CHECK: if (1 && 1 - 0 && (&MemRef_B[c + 1] <= &MemRef_A[0] || &MemRef_A[1024] <= &MemRef_B[c])) ; CHECK: for (int c0 = 0; c0 <= 1023; c0 += 1) ; CHECK: Stmt_for_body(c0); ; CHECK: else Index: test/Isl/Ast/aliasing_parametric_simple_2.ll =================================================================== --- test/Isl/Ast/aliasing_parametric_simple_2.ll +++ test/Isl/Ast/aliasing_parametric_simple_2.ll @@ -5,7 +5,7 @@ ; A[i] = B[c - 10] + B[5]; ; } ; -; CHECK: if (c >= -{{[0-9]*}} && (&MemRef_B[c <= 15 ? 6 : c - 9] <= &MemRef_A[0] || &MemRef_A[1024] <= &MemRef_B[c >= 15 ? 5 : c - 10])) +; CHECK: if (1 && 1 - (c <= -{{[0-9]*}}) && (&MemRef_B[c <= 15 ? 6 : c - 9] <= &MemRef_A[0] || &MemRef_A[1024] <= &MemRef_B[c >= 15 ? 5 : c - 10])) ; CHECK: for (int c0 = 0; c0 <= 1023; c0 += 1) ; CHECK: Stmt_for_body(c0); ; CHECK: else Index: test/Isl/Ast/run-time-condition.ll =================================================================== --- test/Isl/Ast/run-time-condition.ll +++ test/Isl/Ast/run-time-condition.ll @@ -27,7 +27,7 @@ ret void } -; CHECK: if (1) +; CHECK: if (1 && 1 - 0) ; CHECK: for (int c0 = 0; c0 <= 1023; c0 += 1) ; CHECK: Stmt_for_body(c0); ; CHECK: else Index: test/Isl/Ast/simple-run-time-condition.ll =================================================================== --- test/Isl/Ast/simple-run-time-condition.ll +++ test/Isl/Ast/simple-run-time-condition.ll @@ -17,11 +17,7 @@ ; for the delinearization is simplified such that conditions that would not ; cause any code to be executed are not generated. -; CHECK: if ( -; CHECK: (o >= 1 && n + p <= 9223372036854775808 && q <= 0 && m + q >= 0) -; CHECK: || -; CHECK; (o <= 0 && m + q >= 100 && q <= 100) -; CHECK: ) +; CHECK: if (((o >= 1 && q <= 0 && m + q >= 0) || (o <= 0 && m + q >= 100 && q <= 100)) && 1 - (n + p >= 9223372036854775809 || (o <= 0 && m + q >= 9223372036854775909) || (o <= 0 && q <= -9223372036854775709))) ; CHECK: if (o <= 0) { ; CHECK: for (int c0 = 0; c0 < n; c0 += 1) Index: test/Isl/CodeGen/aliasing_parametric_simple_2.ll =================================================================== --- test/Isl/CodeGen/aliasing_parametric_simple_2.ll +++ test/Isl/CodeGen/aliasing_parametric_simple_2.ll @@ -7,6 +7,7 @@ ; ; CHECK: sext i32 %c to i64 ; CHECK: sext i32 %c to i64 +; CHECK: %[[Ctx:[._a-zA-Z0-9]*]] = and i1 true ; CHECK: %[[M0:[._a-zA-Z0-9]*]] = sext i32 %c to i64 ; CHECK: %[[M1:[._a-zA-Z0-9]*]] = icmp sle i64 %[[M0]], 15 ; CHECK: %[[M2:[._a-zA-Z0-9]*]] = sext i32 %c to i64 @@ -24,7 +25,7 @@ ; CHECK: %[[BMin:[._a-zA-Z0-9]*]] = getelementptr i32, i32* %B, i64 %[[m4]] ; CHECK: %[[AltB:[._a-zA-Z0-9]*]] = icmp ule i32* %[[AMax]], %[[BMin]] ; CHECK: %[[NoAlias:[._a-zA-Z0-9]*]] = or i1 %[[BltA]], %[[AltB]] -; CHECK: %[[RTC:[._a-zA-Z0-9]*]] = and i1 %3, %[[NoAlias]] +; CHECK: %[[RTC:[._a-zA-Z0-9]*]] = and i1 %[[Ctx]], %[[NoAlias]] ; CHECK: br i1 %[[RTC]], label %polly.start, label %for.cond ; target datalayout = "e-m:e-i64:64-f80:128-n8:16:32:64-S128" Index: test/Isl/CodeGen/exprModDiv.ll =================================================================== --- test/Isl/CodeGen/exprModDiv.ll +++ test/Isl/CodeGen/exprModDiv.ll @@ -37,9 +37,9 @@ ; CHECK: %pexp.fdiv_q.2 = icmp slt i64 %p, 0 ; CHECK: %pexp.fdiv_q.3 = select i1 %pexp.fdiv_q.2, i64 %pexp.fdiv_q.1, i64 %p ; CHECK: %pexp.fdiv_q.4 = sdiv i64 %pexp.fdiv_q.3, 127 -; CHECK: %17 = mul nsw i64 127, %pexp.fdiv_q.4 -; CHECK: %18 = sub nsw i64 %p, %17 -; CHECK: %polly.access.A8 = getelementptr float, float* %A, i64 %18 +; CHECK: %[[r1:[0-9]*]] = mul nsw i64 127, %pexp.fdiv_q.4 +; CHECK: %[[r2:[0-9]*]] = sub nsw i64 %p, %[[r1]] +; CHECK: %polly.access.A8 = getelementptr float, float* %A, i64 %[[r2]] ; A[p / 127] ; CHECK: %pexp.div = sdiv exact i64 %p, 127 @@ -56,9 +56,9 @@ ; #define floord(n,d) ((n < 0) ? (n - d + 1) : n) / d ; A[p + 128 * floord(-p - 1, 128) + 128] ; POW2: %polly.fdiv_q.shr = ashr i64 %p, 7 -; POW2: %17 = mul nsw i64 128, %polly.fdiv_q.shr -; POW2: %18 = sub nsw i64 %p, %17 -; POW2: %polly.access.A8 = getelementptr float, float* %A, i64 %18 +; POW2: %[[r1:[0-9]*]] = mul nsw i64 128, %polly.fdiv_q.shr +; POW2: %[[r2:[0-9]*]] = sub nsw i64 %p, %[[r1]] +; POW2: %polly.access.A8 = getelementptr float, float* %A, i64 %[[r2]] ; A[p / 128] ; POW2: %pexp.div = sdiv exact i64 %p, 128 Index: test/Isl/CodeGen/large-numbers-in-boundary-context.ll =================================================================== --- test/Isl/CodeGen/large-numbers-in-boundary-context.ll +++ test/Isl/CodeGen/large-numbers-in-boundary-context.ll @@ -4,7 +4,7 @@ ; we will check that we use an appropriaty typed constant, here with 65 bits. ; An alternative would be to bail out early but that would not be as easy. ; -; CHECK: {{.*}} = icmp sge i65 {{.*}}, -9223372036854775809 +; CHECK: {{.*}} = icmp sle i65 {{.*}}, -9223372036854775810 ; ; CHECK: polly.start ; Index: test/Isl/CodeGen/two-scops-in-row.ll =================================================================== --- test/Isl/CodeGen/two-scops-in-row.ll +++ test/Isl/CodeGen/two-scops-in-row.ll @@ -10,7 +10,7 @@ ; SCALAR: Stmt_for_1(c0); ; SCALAR: } -; SCALAR: if (1) +; SCALAR: if (1 && 1 - 0) ; SCALAR: Stmt_for_0(0); Index: test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_1.ll =================================================================== --- test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_1.ll +++ test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_1.ll @@ -16,8 +16,8 @@ ; SCALAR-NEXT: { : } ; SCALAR-NEXT: Assumed Context: ; SCALAR-NEXT: { : } -; SCALAR-NEXT: Boundary Context: -; SCALAR-NEXT: { : } +; SCALAR-NEXT: Invalid Context: +; SCALAR-NEXT: { : 1 = 0 } ; SCALAR-NEXT: Arrays { ; SCALAR-NEXT: i32 MemRef_C[*]; // Element size 4 ; SCALAR-NEXT: i32 MemRef_A[*]; // Element size 4 Index: test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_2.ll =================================================================== --- test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_2.ll +++ test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_2.ll @@ -25,7 +25,7 @@ ; 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: [p_0, p_1, p_2] -> { : 1 = 0 } ; INNERMOST-NEXT: p0: {0,+,{0,+,1}<%bb11>}<%bb13> ; INNERMOST-NEXT: p1: {0,+,1}<%bb11> ; INNERMOST-NEXT: p2: {0,+,1}<%bb13> @@ -75,8 +75,8 @@ ; ALL-NEXT: { : } ; ALL-NEXT: Assumed Context: ; ALL-NEXT: { : } -; ALL-NEXT: Boundary Context: -; ALL-NEXT: { : } +; ALL-NEXT: Invalid Context: +; ALL-NEXT: { : 1 = 0 } ; ALL-NEXT: Arrays { ; ALL-NEXT: i32 MemRef_A[*]; // Element size 4 ; ALL-NEXT: } Index: test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_3.ll =================================================================== --- test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_3.ll +++ test/ScopInfo/NonAffine/non-affine-loop-condition-dependent-access_3.ll @@ -23,8 +23,8 @@ ; 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: Invalid Context: +; INNERMOST-NEXT: [p_0, p_1, p_2] -> { : 1 = 0 } ; INNERMOST-NEXT: p0: {0,+,{0,+,1}<%bb11>}<%bb13> ; INNERMOST-NEXT: p1: {0,+,1}<%bb11> ; INNERMOST-NEXT: p2: {0,+,1}<%bb13> @@ -74,8 +74,8 @@ ; ALL-NEXT: { : } ; ALL-NEXT: Assumed Context: ; ALL-NEXT: { : } -; ALL-NEXT: Boundary Context: -; ALL-NEXT: { : } +; ALL-NEXT: Invalid Context: +; ALL-NEXT: { : 1 = 0 } ; ALL-NEXT: Arrays { ; ALL-NEXT: i32 MemRef_A[*]; // Element size 4 ; ALL-NEXT: } Index: test/ScopInfo/NonAffine/non_affine_conditional_surrounding_affine_loop.ll =================================================================== --- test/ScopInfo/NonAffine/non_affine_conditional_surrounding_affine_loop.ll +++ test/ScopInfo/NonAffine/non_affine_conditional_surrounding_affine_loop.ll @@ -29,8 +29,8 @@ ; ALL-NEXT: { : } ; ALL-NEXT: Assumed Context: ; ALL-NEXT: { : } -; ALL-NEXT: Boundary Context: -; ALL-NEXT: { : } +; ALL-NEXT: Invalid Context: +; ALL-NEXT: { : 1 = 0 } ; ALL-NEXT: Arrays { ; ALL-NEXT: i32 MemRef_A[*]; // Element size 4 ; ALL-NEXT: } Index: test/ScopInfo/NonAffine/non_affine_conditional_surrounding_non_affine_loop.ll =================================================================== --- test/ScopInfo/NonAffine/non_affine_conditional_surrounding_non_affine_loop.ll +++ test/ScopInfo/NonAffine/non_affine_conditional_surrounding_non_affine_loop.ll @@ -33,8 +33,8 @@ ; ALL-NEXT: { : } ; ALL-NEXT: Assumed Context: ; ALL-NEXT: { : } -; ALL-NEXT: Boundary Context: -; ALL-NEXT: { : } +; ALL-NEXT: Invalid Context: +; ALL-NEXT: { : 1 = 0 } ; ALL-NEXT: Arrays { ; ALL-NEXT: i32 MemRef_A[*]; // Element size 4 ; ALL-NEXT: } Index: test/ScopInfo/NonAffine/non_affine_float_compare.ll =================================================================== --- test/ScopInfo/NonAffine/non_affine_float_compare.ll +++ test/ScopInfo/NonAffine/non_affine_float_compare.ll @@ -16,8 +16,8 @@ ; CHECK-NEXT: { : } ; CHECK-NEXT: Assumed Context: ; CHECK-NEXT: { : } -; CHECK-NEXT: Boundary Context: -; CHECK-NEXT: { : } +; CHECK-NEXT: Invalid Context: +; CHECK-NEXT: { : 1 = 0 } ; CHECK-NEXT: Arrays { ; CHECK-NEXT: float MemRef_A[*]; // Element size 4 ; CHECK-NEXT: } Index: test/ScopInfo/NonAffine/non_affine_loop_condition.ll =================================================================== --- test/ScopInfo/NonAffine/non_affine_loop_condition.ll +++ test/ScopInfo/NonAffine/non_affine_loop_condition.ll @@ -30,8 +30,8 @@ ; CHECK-NEXT: { : } ; CHECK-NEXT: Assumed Context: ; CHECK-NEXT: { : } -; CHECK-NEXT: Boundary Context: -; CHECK-NEXT: { : } +; CHECK-NEXT: Invalid Context: +; CHECK-NEXT: { : 1 = 0 } ; CHECK-NEXT: Arrays { ; CHECK-NEXT: i32 MemRef_C[*]; // Element size 4 ; CHECK-NEXT: i32 MemRef_A[*]; // Element size 4 Index: test/ScopInfo/NonAffine/non_affine_loop_used_later.ll =================================================================== --- test/ScopInfo/NonAffine/non_affine_loop_used_later.ll +++ test/ScopInfo/NonAffine/non_affine_loop_used_later.ll @@ -18,8 +18,8 @@ ; CHECK-NEXT: [N] -> { : -2147483648 <= N <= 2147483647 } ; CHECK-NEXT: Assumed Context: ; CHECK-NEXT: [N] -> { : } -; CHECK-NEXT: Boundary Context: -; CHECK-NEXT: [N] -> { : } +; CHECK-NEXT: Invalid Context: +; CHECK-NEXT: [N] -> { : 1 = 0 } ; CHECK-NEXT: p0: %N ; CHECK-NEXT: Arrays { ; CHECK-NEXT: i32 MemRef_j_0__phi; // Element size 4 Index: test/ScopInfo/assume_gep_bounds_2.ll =================================================================== --- test/ScopInfo/assume_gep_bounds_2.ll +++ test/ScopInfo/assume_gep_bounds_2.ll @@ -16,7 +16,7 @@ ; accessed. In this case the value of m does not matter. ; CHECK: Assumed Context: -; CHECK-NEXT: [n, m, p] -> { : p <= 20 and (n <= 0 or (n > 0 and m <= 20)) } +; CHECK-NEXT: [n, m, p] -> { : p <= 20 and (n <= 0 or (n > 0 and m <= 20)) } target datalayout = "e-m:e-i64:64-f80:128-n8:16:32:64-S128" Index: test/ScopInfo/bounded_loop_assumptions.ll =================================================================== --- test/ScopInfo/bounded_loop_assumptions.ll +++ test/ScopInfo/bounded_loop_assumptions.ll @@ -7,7 +7,9 @@ ; finally, if N == 2 we would have an unbounded inner loop. ; ; CHECK: Assumed Context: -; CHECK-NEXT: [N] -> { : N >= 3 or N <= 1 } +; CHECK-NEXT: [N] -> { : } +; CHECK: Invalid Context: +; CHECK-NEXT: [N] -> { : N = 2 } ; ; int jd(int *restrict A, int x, int N) { ; for (int i = 1; i < N; i++) Index: test/ScopInfo/long-sequence-of-error-blocks-2.ll =================================================================== --- test/ScopInfo/long-sequence-of-error-blocks-2.ll +++ test/ScopInfo/long-sequence-of-error-blocks-2.ll @@ -6,11 +6,13 @@ %struct.hoge = type { i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [8 x [2 x i32]], [8 x [2 x i32]], [4 x [4 x i32]], i32, i32, i32, i32, [256 x i8], [256 x i8], [256 x i8], [256 x i8], [256 x i8], i32, i32, i32, i32, i32, i32, [500 x i8], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [256 x i8], [256 x i8], [256 x i8], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [1024 x i8], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, double, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [256 x i8], [256 x i8], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [256 x i8], i32, i32, i32*, i32*, i8*, i32*, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, double, double, double, [5 x double], i32, [8 x i32], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [6 x double], [6 x double], [256 x i8], i32, i32, i32, i32, [2 x [5 x i32]], [2 x [5 x i32]], i32, i32, i32, i32, i32, i32, i32, i32, i32, [3 x i32], i32 } ; The assumed context of this test case has at some point become very complex. -; This test case verifies that we bail out and generate an infeasible -; assumed context. It should be updated after having made the assumed context -; construction more efficient. - -; CHECK-NOT: Assumed Context: +; However, since we keep both the assumed as well as invalid context that +; problem is solved. +; +; CHECK: Assumed Context: +; CHECK-NEXT: [tmp17, tmp21, tmp27, tmp31, tmp37, tmp41, tmp46, tmp52, tmp56, tmp62] -> { : } +; CHECK: Invalid Context: +; CHECK-NEXT: [tmp17, tmp21, tmp27, tmp31, tmp37, tmp41, tmp46, tmp52, tmp56, tmp62] -> { : (tmp17 < 0 and tmp21 < 0) or (tmp17 < 0 and tmp21 > 0) or (tmp17 > 0 and tmp21 < 0) or (tmp17 > 0 and tmp21 > 0) or (tmp37 < 0 and tmp41 < 0 and tmp46 > 0) or (tmp37 < 0 and tmp41 > 0 and tmp46 > 0) or (tmp37 > 0 and tmp41 < 0 and tmp46 > 0) or (tmp37 > 0 and tmp41 > 0 and tmp46 > 0) or (tmp27 = 3 and tmp31 <= 143) or (tmp56 = 0 and tmp52 < 0) or (tmp56 = 0 and tmp52 > 0) } @global = external global [300 x i8], align 16 Index: test/ScopInfo/long-sequence-of-error-blocks.ll =================================================================== --- test/ScopInfo/long-sequence-of-error-blocks.ll +++ test/ScopInfo/long-sequence-of-error-blocks.ll @@ -5,15 +5,14 @@ %struct.hoge = type { i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [8 x [2 x i32]], [8 x [2 x i32]], [4 x [4 x i32]], i32, i32, i32, i32, [256 x i8], [256 x i8], [256 x i8], [256 x i8], [256 x i8], i32, i32, i32, i32, i32, i32, [500 x i8], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [256 x i8], [256 x i8], [256 x i8], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [1024 x i8], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, double, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [256 x i8], [256 x i8], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [256 x i8], i32, i32, i32*, i32*, i8*, i32*, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, double, double, double, [5 x double], i32, [8 x i32], i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, [6 x double], [6 x double], [256 x i8], i32, i32, i32, i32, [2 x [5 x i32]], [2 x [5 x i32]], i32, i32, i32, i32, i32, i32, i32, i32, i32, [3 x i32], i32 } -; The assumed context of this test case is still very large and should be -; reduced. - ; This test case contains a long sequence of branch instructions together with ; function calls that are considered 'error blocks'. We verify that the ; iteration spaces are not overly complicated. ; CHECK: Assumed Context: -; CHECK-NEXT: [tmp17, tmp21, tmp27, tmp31] -> { : (tmp17 = 0 and tmp27 >= 4) or (tmp17 = 0 and tmp27 <= 2) or (tmp21 = 0 and tmp17 < 0 and tmp27 >= 4) or (tmp21 = 0 and tmp17 < 0 and tmp27 <= 2) or (tmp21 = 0 and tmp17 > 0 and tmp27 >= 4) or (tmp21 = 0 and tmp17 > 0 and tmp27 <= 2) or (tmp17 = 0 and tmp27 = 3 and tmp31 >= 144) or (tmp21 = 0 and tmp27 = 3 and tmp17 < 0 and tmp31 >= 144) or (tmp21 = 0 and tmp27 = 3 and tmp17 > 0 and tmp31 >= 144) } +; CHECK-NEXT: [tmp17, tmp21, tmp27, tmp31] -> { : } +; CHECK: Invalid Context: +; CHECK-NEXT: [tmp17, tmp21, tmp27, tmp31] -> { : (tmp17 < 0 and tmp21 < 0) or (tmp17 < 0 and tmp21 > 0) or (tmp17 > 0 and tmp21 < 0) or (tmp17 > 0 and tmp21 > 0) or (tmp27 = 3 and tmp31 <= 143) } ; ; CHECK: Statements { ; CHECK-NEXT: Stmt_bb15 Index: test/ScopInfo/multidim_3d_parametric_array_static_loop_bounds.ll =================================================================== --- test/ScopInfo/multidim_3d_parametric_array_static_loop_bounds.ll +++ test/ScopInfo/multidim_3d_parametric_array_static_loop_bounds.ll @@ -11,6 +11,8 @@ ; CHECK: Assumed Context: ; CHECK-NEXT: [m, o] -> { : m >= 150 and o >= 200 } +; CHECK: Invalid Context: +; CHECK-NEXT: [m, o] -> { : 1 = 0 } ; ; CHECK: p0: %m ; CHECK-NEXT: p1: %o Index: test/ScopInfo/non-pure-function-call.ll =================================================================== --- test/ScopInfo/non-pure-function-call.ll +++ test/ScopInfo/non-pure-function-call.ll @@ -1,7 +1,9 @@ ; RUN: opt %loadPolly -polly-scops -analyze < %s | FileCheck %s ; ; CHECK: Assumed Context: -; CHECK-NEXT: [N] -> { : N <= 101 } +; CHECK-NEXT: [N] -> { : } +; CHECK: Invalid Context: +; CHECK-NEXT: [N] -> { : N >= 102 } ; ; void g(void); ; void f(int *A, int N) { Index: test/ScopInfo/non-pure-function-calls-causes-dead-blocks.ll =================================================================== --- test/ScopInfo/non-pure-function-calls-causes-dead-blocks.ll +++ test/ScopInfo/non-pure-function-calls-causes-dead-blocks.ll @@ -35,7 +35,9 @@ ; ; CHECK: Region: %entry.split---%if.end.20 ; CHECK: Assumed Context: -; CHECK-NEXT: [timeit, N] -> { : timeit = 0 } +; CHECK-NEXT: [timeit, N] -> { : } +; CHECK: Invalid Context: +; CHECK-NEXT: [timeit, N] -> { : timeit < 0 or timeit > 0 } ; CHECK: Statements { ; CHECK-NOT: Stmt_if_then_split ; CHECK: Stmt_for_body Index: test/ScopInfo/non-pure-function-calls.ll =================================================================== --- test/ScopInfo/non-pure-function-calls.ll +++ test/ScopInfo/non-pure-function-calls.ll @@ -25,14 +25,16 @@ ; timer_stop(); ; } ; -; CHECK: Region: %for.cond---%if.end.20 -; CHECK: Assumed Context: -; CHECK: [N, timeit] -> { : timeit = 0 } -; CHECK: Statements { -; CHECK: Stmt -; CHECK: Stmt -; CHECK-NOT Stmt -; CHECK: } +; CHECK: Region: %for.cond---%if.end.20 +; CHECK: Assumed Context: +; CHECK-NEXT: [N, timeit] -> { : } +; CHECK: Invalid Context: +; CHECK-NEXT: [N, timeit] -> { : timeit < 0 or timeit > 0 } +; CHECK: Statements { +; CHECK: Stmt +; CHECK: Stmt +; CHECK-NOT: Stmt +; CHECK: } ; target datalayout = "e-m:e-i64:64-f80:128-n8:16:32:64-S128" Index: test/ScopInfo/remarks.ll =================================================================== --- test/ScopInfo/remarks.ll +++ test/ScopInfo/remarks.ll @@ -1,10 +1,10 @@ ; RUN: opt %loadPolly -pass-remarks-analysis="polly-scops" -polly-scops -disable-output < %s 2>&1 | FileCheck %s ; ; CHECK: remark: test/ScopInfo/remarks.c:4:7: SCoP begins here. -; CHECK: remark: test/ScopInfo/remarks.c:8:5: Finite loop assumption: [M, N, Debug] -> { : M >= 0 or (M <= -2 and N <= 0) or (M = -1 and N <= 0) } -; CHECK: remark: test/ScopInfo/remarks.c:13:7: No-error assumption: [M, N, Debug] -> { : M < 0 or (M >= 0 and N <= 0) or (Debug = 0 and M >= 0 and N > 0) } -; CHECK: remark: test/ScopInfo/remarks.c:9:7: Inbounds assumption: [M, N, Debug] -> { : M <= 100 or (M > 0 and N <= 0) } -; CHECK: remark: :0:0: No-overflows assumption: [N, M, Debug] -> { : -2147483648 - N <= M <= 2147483647 - N } +; CHECK: remark: test/ScopInfo/remarks.c:8:5: Finite loop restriction: [M, N, Debug] -> { : N > 0 and (M <= -2 or M = -1) } +; CHECK: remark: test/ScopInfo/remarks.c:13:7: No-error restriction: [M, N, Debug] -> { : M >= 0 and N > 0 and (Debug < 0 or Debug > 0) } +; CHECK: remark: test/ScopInfo/remarks.c:9:7: Inbounds assumption: [M, N, Debug] -> { : M <= 100 or (M > 0 and N <= 0) } +; CHECK: remark: :0:0: No-overflows restriction: [N, M, Debug] -> { : M <= -2147483649 - N or M >= 2147483648 - N } ; CHECK: remark: test/ScopInfo/remarks.c:9:18: Possibly aliasing pointer, use restrict keyword. ; CHECK: remark: test/ScopInfo/remarks.c:9:33: Possibly aliasing pointer, use restrict keyword. ; CHECK: remark: test/ScopInfo/remarks.c:9:15: Possibly aliasing pointer, use restrict keyword. Index: test/ScopInfo/switch-1.ll =================================================================== --- test/ScopInfo/switch-1.ll +++ test/ScopInfo/switch-1.ll @@ -48,7 +48,7 @@ ; CHECK-NEXT: [N] -> { Stmt_sw_bb_6[i0] -> MemRef_A[i0] }; ; CHECK-NEXT: } -; AST: if (1) +; AST: if (1 && 1 - 0) ; ; AST: { ; AST-NEXT: for (int c0 = 1; c0 < N - 2; c0 += 4) { Index: test/ScopInfo/switch-2.ll =================================================================== --- test/ScopInfo/switch-2.ll +++ test/ScopInfo/switch-2.ll @@ -38,7 +38,7 @@ ; CHECK-NEXT: [N] -> { Stmt_sw_bb_2[i0] -> MemRef_A[i0] }; ; CHECK-NEXT: } -; AST: if (1) +; AST: if (1 && 1 - 0) ; ; AST: for (int c0 = 0; c0 < N; c0 += 4) { ; AST-NEXT: Stmt_sw_bb(c0); Index: test/ScopInfo/switch-3.ll =================================================================== --- test/ScopInfo/switch-3.ll +++ test/ScopInfo/switch-3.ll @@ -56,7 +56,7 @@ ; CHECK-NEXT: [N] -> { Stmt_sw_bb_9[i0] -> MemRef_A[i0] }; ; CHECK-NEXT: } -; AST: if (1) +; AST: if (1 && 1 - 0) ; ; AST: for (int c0 = 0; c0 < N; c0 += 1) { ; AST-NEXT: if ((c0 - 2) % 4 == 0) Index: test/ScopInfo/switch-4.ll =================================================================== --- test/ScopInfo/switch-4.ll +++ test/ScopInfo/switch-4.ll @@ -60,7 +60,7 @@ ; CHECK-NEXT: [N] -> { Stmt_sw_bb_9[i0] -> MemRef_A[i0] }; ; CHECK-NEXT: } -; AST: if (1) +; AST: if (1 && 1 - 0) ; ; AST: { ; AST-NEXT: for (int c0 = 0; c0 < N - 3; c0 += 4) { Index: test/ScopInfo/switch-5.ll =================================================================== --- test/ScopInfo/switch-5.ll +++ test/ScopInfo/switch-5.ll @@ -26,7 +26,7 @@ ; CHECK-NEXT: [N] -> { Stmt_for_body[i0] -> MemRef_A[0] }; ; CHECK-NEXT: } -; AST: if (1) +; AST: if (1 && 1 - 0) ; ; AST: if (N >= 1) ; AST-NEXT: Stmt_for_body(0); Index: test/ScopInfo/switch-6.ll =================================================================== --- test/ScopInfo/switch-6.ll +++ test/ScopInfo/switch-6.ll @@ -60,7 +60,7 @@ ; CHECK-NEXT: [N] -> { Stmt_sw_bb_9[i0] -> MemRef_A[3] }; ; CHECK-NEXT: } -; AST: if (1) +; AST: if (1 && 1 - 0) ; ; AST: if (N >= 1) { ; AST-NEXT: Stmt_sw_bb(0); Index: test/ScopInfo/switch-7.ll =================================================================== --- test/ScopInfo/switch-7.ll +++ test/ScopInfo/switch-7.ll @@ -42,7 +42,7 @@ ; CHECK-NEXT: [c, N] -> { Stmt_for_body_7[i0] -> MemRef_A[1 + i0] }; ; CHECK-NEXT: } -; AST: if (1) +; AST: if (1 && 1 - 0) ; ; AST: if (c == 1) { ; AST-NEXT: for (int c0 = 0; c0 < N; c0 += 1) Index: test/ScopInfo/test-wrapping-in-condition.ll =================================================================== --- test/ScopInfo/test-wrapping-in-condition.ll +++ test/ScopInfo/test-wrapping-in-condition.ll @@ -1,7 +1,7 @@ ; RUN: opt %loadPolly -polly-scops -analyze < %s | FileCheck %s ; -; CHECK: Boundary Context: -; CHECK: [N] -> { : N <= 128 } +; CHECK: Invalid Context: +; CHECK: [N] -> { : N >= 129 } ; ; #include ; #include Index: test/ScopInfo/user_provided_assumptions.ll =================================================================== --- test/ScopInfo/user_provided_assumptions.ll +++ test/ScopInfo/user_provided_assumptions.ll @@ -12,8 +12,8 @@ ; SCOP-NEXT: [N, M, Debug] -> { : Debug = 0 and N > 0 and 0 < M <= 2147483647 - N and M <= 100 } ; SCOP: Assumed Context: ; SCOP-NEXT: [N, M, Debug] -> { : } -; SCOP: Boundary Context: -; SCOP-NEXT: [N, M, Debug] -> { : } +; SCOP: Invalid Context: +; SCOP-NEXT: [N, M, Debug] -> { : 1 = 0 } ; ; #include ; Index: test/ScopInfo/wraping_signed_expr_0.ll =================================================================== --- test/ScopInfo/wraping_signed_expr_0.ll +++ test/ScopInfo/wraping_signed_expr_0.ll @@ -10,8 +10,8 @@ ; we will add the assumption that i+1 won't overflow only to the former. ; ; CHECK: Function: wrap -; CHECK: Boundary Context: -; CHECK: [N] -> { : N <= 125 } +; CHECK: Invalid Context: +; CHECK: [N] -> { : N >= 126 } ; ; ; FIXME: This is a negative test as nowrap should not need an assumed context. @@ -19,7 +19,7 @@ ; which lacks the flags we would need to avoid runtime checks. ; ; CHECK: Function: nowrap -; CHECK: Boundary Context: +; CHECK: Invalid Context: ; CHECK-NOT: [N] -> { : } ; target datalayout = "e-m:e-i8:64-f80:128-n8:16:32:64-S128" Index: test/ScopInfo/wraping_signed_expr_1.ll =================================================================== --- test/ScopInfo/wraping_signed_expr_1.ll +++ test/ScopInfo/wraping_signed_expr_1.ll @@ -16,12 +16,12 @@ ; sizeof(long) == 8 ; ; CHECK: Function: wrap -; CHECK: Boundary Context: -; CHECK: [N] -> { : N <= 1152921504606846975 } +; CHECK: Invalid Context: +; CHECK-NEXT: [N] -> { : N >= 1152921504606846976 } ; ; CHECK: Function: nowrap -; CHECK: Boundary Context: -; CHECK: [N] -> { : } +; CHECK: Invalid Context: +; CHECK-NEXT: [N] -> { : 1 = 0 } ; target datalayout = "e-m:e-i64:64-f80:128-n8:16:32:64-S128" Index: test/ScopInfo/wraping_signed_expr_2.ll =================================================================== --- test/ScopInfo/wraping_signed_expr_2.ll +++ test/ScopInfo/wraping_signed_expr_2.ll @@ -15,8 +15,8 @@ ; CHECK: Context: ; CHECK-NEXT: [N] -> { : -2147483648 <= N <= 2147483647 } ; -; CHECK: Boundary Context: -; CHECK-NEXT: [N] -> { : N <= 2147483618 } +; CHECK: Invalid Context: +; CHECK-NEXT: [N] -> { : N >= 2147483619 } target datalayout = "e-m:e-i32:64-f80:128-n8:16:32:64-S128" Index: test/ScopInfo/wraping_signed_expr_3.ll =================================================================== --- test/ScopInfo/wraping_signed_expr_3.ll +++ test/ScopInfo/wraping_signed_expr_3.ll @@ -8,8 +8,8 @@ ; Note: 2147483648 == 2 ^ 31 ; ; CHECK: Function: wrap -; CHECK: Boundary Context: -; CHECK: [N, p] -> { : p <= 2147483648 - N } +; CHECK: Invalid Context: +; CHECK: [N, p] -> { : p >= 2147483649 - N } ; target datalayout = "e-m:e-i32:64-f80:128-n8:16:32:64-S128" Index: test/ScopInfo/wraping_signed_expr_4.ll =================================================================== --- test/ScopInfo/wraping_signed_expr_4.ll +++ test/ScopInfo/wraping_signed_expr_4.ll @@ -10,8 +10,8 @@ ; CHECK: Context: ; CHECK-NEXT: [N, p] -> { : -128 <= N <= 127 and -128 <= p <= 127 } ; -; CHECK: Boundary Context: -; CHECK-NEXT: [N, p] -> { : p >= -127 } +; CHECK: Invalid Context: +; CHECK-NEXT: [N, p] -> { : p = -128 } target datalayout = "e-m:e-i8:64-f80:128-n8:16:32:64-S128" Index: test/ScopInfo/wraping_signed_expr_5.ll =================================================================== --- test/ScopInfo/wraping_signed_expr_5.ll +++ test/ScopInfo/wraping_signed_expr_5.ll @@ -3,8 +3,8 @@ ; We should not generate runtime check for ((int)r1 + (int)r2) as it is known not ; to overflow. However (p + q) can, thus checks are needed. ; -; CHECK: Boundary Context: -; CHECK-NEXT: [r1, r2, q, p] -> { : r2 <= 127 + r1 and -2147483648 - q <= p <= 2147483647 - q } +; CHECK: Invalid Context: +; CHECK-NEXT: [r1, r2, q, p] -> { : p <= -2147483649 - q or p >= 2147483648 - q or r2 >= 128 + r1 } ; ; void wraps(int *A, int p, short q, char r1, char r2) { ; for (char i = r1; i < r2; i++) Index: test/ScopInfo/wraping_signed_expr_6.ll =================================================================== --- test/ScopInfo/wraping_signed_expr_6.ll +++ test/ScopInfo/wraping_signed_expr_6.ll @@ -1,7 +1,7 @@ ; RUN: opt %loadPolly -polly-scops -analyze < %s | FileCheck %s ; -; CHECK: Boundary Context: -; CHECK: [N] -> { : N <= 128 } +; CHECK: Invalid Context: +; CHECK: [N] -> { : N >= 129 } ; ; void foo(float *A, long N) { ; for (long i = 0; i < N; i++) Index: test/ScopInfo/wraping_signed_expr_7.ll =================================================================== --- test/ScopInfo/wraping_signed_expr_7.ll +++ test/ScopInfo/wraping_signed_expr_7.ll @@ -1,7 +1,7 @@ ; RUN: opt %loadPolly -polly-scops -analyze < %s | FileCheck %s ; -; CHECK: Boundary Context: -; CHECK: [N] -> { : N <= 128 } +; CHECK: Invalid Context: +; CHECK: [N] -> { : N >= 129 } ; ; void foo(float *A, long N) { ; for (long i = 0; i < N;)