summaryrefslogtreecommitdiffstats
path: root/polly/lib
Commit message (Collapse)AuthorAgeFilesLines
...
* [BlockGenerator] Take context into account when identifying partial writesTobias Grosser2017-06-011-1/+4
| | | | | | | | | | | | | | | | | | | | | | | | | | A partial write is a write where the domain of the values written is a subset of the execution domain of the parent statement containing the write. Originally, we directly checked this subset relation whereas it is indeed only important that the subset relation holds for the parameter values that are known to be valid in the execution context of the scop. We update our check to avoid the unnecessary introduction of partial writes in situations where the write appears to be partial without context information, but where context information allows us to understand that a full write can be generated. This change fixes (hides) a recent regression introduced in r303517, which broke our AOSP builds. The part that is correctly fixed in this change is that we do not any more unnecessarily generate a partial write. This is good performance wise and, as we currently do not yet explicitly introduce partial writes in the default configuration, this also hides possible bugs in the partial writes implementation. The crashes that we have originally seen were caused by such a bug, where partial writes were incorrectly generated in region statements. An additional patch in a subsequent commit is needed to address this problem. Reported-by: Reported-by: Eli Friedman <efriedma@codeaurora.org> Differential Revision: https://reviews.llvm.org/D33759 llvm-svn: 304398
* [BlockGenerator] Translate buildContainsCondition to idiomatic isl C++Tobias Grosser2017-05-311-18/+11
| | | | llvm-svn: 304354
* [isl++] Update bindingsTobias Grosser2017-05-311-1/+13
| | | | | | | | This change removes the requirement for explicit conversions from isl::boolean to isl::bool, which resolves a compilation error on OSX. Suggested-by: Siddharth Bhat <siddu.druid@gmail.com> llvm-svn: 304288
* [ScopInfo] Do not add terminator & synthesizable instructions to the output ↵Michael Kruse2017-05-291-2/+5
| | | | | | | | | | | instructions. Such instructions are generates on-demand by the CodeGenerator and thus do not need representation in a statement. Differential Revision: https://reviews.llvm.org/D33642 llvm-svn: 304151
* Revert "[NFC] Fix formatting & typecast issue. Build succeeds."Siddharth Bhat2017-05-292-2/+2
| | | | | | | | Should not have 'fixed' the formatting issue, I did not have the most recent version of `clang-format`. This reverts commit 761b1268359e14e59142f253d77864a29d55c56c. llvm-svn: 304148
* [NFC] Fix formatting & typecast issue. Build succeeds.Siddharth Bhat2017-05-292-2/+2
| | | | | | | | | - Fix formatting in `RegisterPasses.cpp`. - `assert` tried to compare `isl::boolean` against `long`. Explicitly construct `bool` from `isl::boolean`. This allows the implicit cast of `bool` to `long. llvm-svn: 304146
* Adapt to recent clang-format changesTobias Grosser2017-05-291-1/+1
| | | | llvm-svn: 304136
* Delinearize memory accesses that reference parameters coming from function callsTobias Grosser2017-05-273-5/+60
| | | | | | | | | | | | | | Certain affine memory accesses which we model today might contain products of parameters which we might combined into a new parameter to be able to create an affine expression that represents these memory accesses. Especially in the context of OpenCL, this approach looses information as memory accesses such as A[get_global_id(0) * N + get_global_id(1)] are assumed to be linear. We correctly recover their multi-dimensional structure by assuming that parameters that are the result of a function call at IR level likely are not parameters, but indeed induction variables. The resulting access is now A[get_global_id(0)][get_global_id(1)] for an array A[][N]. llvm-svn: 304075
* Allow side-effect free function calls in valid affine SCEVsTobias Grosser2017-05-272-3/+66
| | | | | | | | | | | | | | | Side-effect free function calls with only constant parameters can be easily re-generated and consequently do not prevent us from modeling a SCEV. This change allows array subscripts to reference function calls such as 'get_global_id()' as used in OpenCL. We use the function name plus the constant operands to name the parameter. This is possible as the function name is required and is not dropped in release builds the same way names of llvm::Values are dropped. We also provide more readable names for common OpenCL functions, to make it easy to understand the polyhedral model we generate. llvm-svn: 304074
* Update isl to isl-0.18-662-g17e172eTobias Grosser2017-05-2767-812/+2673
| | | | | | This is a general maintenance update llvm-svn: 304069
* [Polly] Added the list of Instructions to output in ScopInfo passTobias Grosser2017-05-272-5/+28
| | | | | | | | | | | | Summary: This patch outputs all the list of instructions in BlockStmts. Reviewers: Meinersbur, grosser, bollu Subscribers: bollu, llvm-commits, pollydev Differential Revision: https://reviews.llvm.org/D33163 llvm-svn: 304062
* [ScopInfo] Translate mapToDimension to isl C++ [NFC]Tobias Grosser2017-05-261-46/+18
| | | | llvm-svn: 304007
* [ScopInfo] Tighten compute out introduced in r303404Tobias Grosser2017-05-241-1/+1
| | | | | | | It seems we are still spending too much time on rare inputs, which continue to timeout the AOSP buildbot. Let's see if a further reduction is sufficient. llvm-svn: 303807
* [Polly] Add handling of Top Level RegionsPhilip Pfaffe2017-05-244-16/+31
| | | | | | | | | | | | | | | | | | | | | | | | | | | Summary: My goal is to make the newly added `AllowWholeFunctions` options more usable/powerful. The changes to ScopBuilder.cpp are exclusively checks to prevent `Region.getExit()` from being dereferenced, since Top Level Regions (TLRs) don't have an exit block. In ScopDetection's `isValidCFG`, I removed a check that disallowed ReturnInstructions to have return values. This might of course have been intentional, so I would welcome your feedback on this and maybe a small explanation why return values are forbidden. Maybe it can be done but needs more changes elsewhere? The remaining changes in ScopDetection are simply to consider the AllowWholeFunctions option in more places, i.e. allow TLRs when it is set and once again avoid derefererncing `getExit()` if it doesn't exist. Finally, in ScopHelper.cpp I extended `polly::isErrorBlock` to handle regions without exit blocks as well: The original check was if a given BasicBlock dominates all predecessors of the exit block. Therefore I do the same for TLRs by regarding all BasicBlocks terminating with a ReturnInst as predecessors of a "virtual" function exit block. Patch by: Lukas Boehm Reviewers: philip.pfaffe, grosser, Meinersbur Reviewed By: grosser Subscribers: pollydev, llvm-commits, bollu Tags: #polly Differential Revision: https://reviews.llvm.org/D33411 llvm-svn: 303790
* [DeLICM] Partial writes for PHIs.Michael Kruse2017-05-241-1/+7
| | | | | | | | | | | | | | Enable the use for partial writes for PHI write accesses with a switch. This simply skips the test for whether a PHI write would be partial. The analog test for partial value writes also protects for partial reads which we do not support (yet). It is possible to test for partial reads separately such that we could skip the partial write check as well. In case this shows up to be useful, I can implement it as well. Differential Revision: https://reviews.llvm.org/D33487 llvm-svn: 303762
* [JSONImporter] misses checks whether the data it imports makes sense.Michael Kruse2017-05-241-7/+145
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | Without this patch, the JSONImporter did not verify if the data it loads were correct or not (Bug llvm.org/PR32543). I add some checks in the JSONImporter class and some test cases. Here are the checks (and test cases) I added : JSONImporter::importContext - The "context" key does not exist. - The context was not parsed successfully by ISL. - The isl_set has the wrong number of parameters. - The isl_set is not a parameter set. JSONImporter::importSchedule - The "statements" key does not exist. - There is not the right number of statement in the file. - The "schedule" key does not exist. - The schedule was not parsed successfully by ISL. JSONImporter::importAccesses - The "statements" key does not exist. - There is not the right number of statement in the file. - The "accesses" key does not exist. - There is not the right number of memory accesses in the file. - The "relation" key does not exist. - The memory access was not parsed successfully by ISL. JSONImporter::areArraysEqual - The "type" key does not exist. - The "sizes" key does not exist. - The "name" key does not exist. JSONImporter::importArrays /!\ Do not check if there is an key name "arrays" because it is not considered as an error. All checks are already in place or implemented in JSONImporter::areArraysEqual. Contributed-by: Nicolas Bonfante <nicolas.bonfante@insa-lyon.fr> Differential Revision: https://reviews.llvm.org/D32739 llvm-svn: 303759
* [Polly][NewPM] Reenable ScopPassManager unittestPhilip Pfaffe2017-05-231-0/+1
| | | | llvm-svn: 303629
* Post-commit fix of a commentPhilip Pfaffe2017-05-231-1/+1
| | | | llvm-svn: 303628
* [Polly][NewPM] Port CodeGen to the new PMPhilip Pfaffe2017-05-231-189/+206
| | | | | | | | | | | | | | | | Summary: To move CG to the new PM I outlined the various helper that were previously members of the CG class into free static functions. The CG class itself I moved into a header, which is required because we need to include it in `RegisterPasses` eventually. Reviewers: grosser, Meinersbur Reviewed By: grosser Subscribers: pollydev, llvm-commits, sanjoy Tags: #polly Differential Revision: https://reviews.llvm.org/D33423 llvm-svn: 303624
* [Polly][NewPM] Port IslAst to the new ScopPassManagerPhilip Pfaffe2017-05-234-59/+77
| | | | | | | | | | | | | | | | Summary: This patch ports IslAst to the new PM. The change is mostly straightforward. The only major modification required is making IslAst move-only, to correctly manage the isl resources it owns. Reviewers: grosser, Meinersbur Reviewed By: grosser Subscribers: nemanjai, pollydev, llvm-commits Tags: #polly Differential Revision: https://reviews.llvm.org/D33422 llvm-svn: 303622
* [Polly][NewPM] Port DependenceInfo to the new ScopPassManager.Philip Pfaffe2017-05-231-0/+42
| | | | | | | | | | | | | | | | Summary: This patch ports DependenceInfo to the new ScopPassManager. Printing is implemented as a seperate printer pass. Reviewers: grosser, Meinersbur Reviewed By: grosser Subscribers: llvm-commits, pollydev Tags: #polly Differential Revision: https://reviews.llvm.org/D33421 llvm-svn: 303621
* [ScopInfo] Translate foldAccessRelation to isl C++ [NFC]Tobias Grosser2017-05-231-50/+47
| | | | llvm-svn: 303615
* [ScopInfo] Translate buildMemIntrinsicAccessRelation to isl C++ [NFC]Tobias Grosser2017-05-231-16/+16
| | | | llvm-svn: 303612
* [ScopInfo] Translate assumeNoOutOfBound to isl C++ [NFC]Tobias Grosser2017-05-231-26/+19
| | | | llvm-svn: 303611
* [ScopInfo] Translate applyAndSetFAD to isl C++Tobias Grosser2017-05-231-10/+6
| | | | llvm-svn: 303610
* [ScopInfo] Translate isReadOnly to isl C++Tobias Grosser2017-05-231-8/+4
| | | | llvm-svn: 303608
* [ScopInfo] Simplify domains earlyTobias Grosser2017-05-231-0/+2
| | | | | | | | | | | | | This speeds up scop modeling for scops with many redundent existentially quantified constraints. For the attached test case, this change reduces scop modeling time from minutes (hours?) to 0.15 seconds. This change resolves a compilation timeout on the AOSP build. Thanks Eli for reporting _and_ reducing the test case! Reported-by: Eli Friedman <efriedma@codeaurora.org> llvm-svn: 303600
* [CodeGen] Add invalidation of the loop SCEVs after merge block generation.Michael Kruse2017-05-221-0/+12
| | | | | | | | | | | | | The SCEVs of loops surrounding the escape users of a merge blocks are forgotten, so that loop trip counts based on old values can be revoked. This fixes llvm.org//PR32536 Contributed-by: Baranidharan Mohan <mbdharan@gmail.com> Differential Revision: https://reviews.llvm.org/D33195 llvm-svn: 303561
* [CodeGen] Support partial write accesses.Michael Kruse2017-05-214-82/+213
| | | | | | | | | | | | | | | | | | | Allow the BlockGenerator to generate memory writes that are not defined over the complete statement domain, but only over a subset of it. It generates a condition that evaluates to 1 if executing the subdomain, and only then execute the access. Only write accesses are supported. Read accesses would require a PHINode which has a value if the access is not executed. Partial write makes DeLICM able to apply mappings that are not defined over the entire domain (for instance, a branch that leaves a loop with a PHINode in its header; a MemoryKind::PHI write when leaving is never read by its PHI read). Differential Revision: https://reviews.llvm.org/D33255 llvm-svn: 303517
* [ScopInfo] Translate updateDimensionality to isl C++ [NFC]Tobias Grosser2017-05-211-33/+29
| | | | llvm-svn: 303514
* [isl++] add isl_constraint to C++ bindings [NFC]Tobias Grosser2017-05-211-0/+396
| | | | llvm-svn: 303512
* [ScopInfo] Translate wrapConstantDimensions to isl C++ [NFC]Tobias Grosser2017-05-211-26/+22
| | | | llvm-svn: 303511
* [ScopInfo] Translate addRangeBoundsToSet to isl C++ [NFC]Tobias Grosser2017-05-211-22/+23
| | | | llvm-svn: 303510
* [ScheduleOptimizer] Move schedule construction to isl C++ [NFC]Tobias Grosser2017-05-211-20/+16
| | | | llvm-svn: 303508
* [Simplify] Move to isl C++Tobias Grosser2017-05-211-19/+13
| | | | llvm-svn: 303507
* [isl++] Export (almost) all functions from islTobias Grosser2017-05-211-11/+7829
| | | | | | | | | | | | | | | | This commit exports the majority of the isl functions to the isl C++ interface. The official isl C++ bindings still require discussions to define the set of functions that are officially supported. As a result, the officially exported functionality will be rather limited until these discussions conclude and a non-trivial set of isl functions is officially supported through the isl C++ bindings. Starting from this commit we ship with Polly an extended version of the official isl C++ bindings to ensure sufficient functionality is available such that LLVM developers can make efficient use of isl through C++. The practical experience Polly gathers with its bindings will then be used to gradually upstream patches to isl to extend the official bindings. llvm-svn: 303506
* [isl++] Rebase isl C++ bindings on top of 29aee98ceTobias Grosser2017-05-212-314/+503
| | | | | | | | This reduces the diff to the official isl C++ bindings and solves a correctness issue with isl::booleans, where isl_bool_error results were accidentally converted to isl::boolean::true. llvm-svn: 303505
* [isl++] Move isl raw_ostream printers into separate headerTobias Grosser2017-05-215-156/+4
| | | | | | | | Instead of relying on these functions to be part of the isl C++ bindings, we just define this functionality independently. This allows us to use isl C++ bindings that do not contain LLVM specific functionality. llvm-svn: 303503
* [Fortran Support] Materialize outermost dimension for Fortran array.Siddharth Bhat2017-05-193-4/+202
| | | | | | | | | | | | | | | | | | | - We use the outermost dimension of arrays since we need this information to generate GPU transfers. - In general, if we do not know the outermost dimension of the array (because the indexing expression is non-affine, for example) then we simply cannot generate transfer code. - However, for Fortran arrays, we can use the Fortran array representation which stores the dimensions of all arrays. - This patch uses the Fortran array representation to generate code that computes the outermost dimension size. Differential Revision: https://reviews.llvm.org/D32967 llvm-svn: 303429
* [ScopDetection] Allow detection of full functionsTobias Grosser2017-05-191-2/+8
| | | | | | This is useful when only analyzing functions. llvm-svn: 303420
* [ScopInfo] Fix typo in documentationTobias Grosser2017-05-191-5/+6
| | | | llvm-svn: 303405
* [ScopInfo] Gracefully handle long compile timesTobias Grosser2017-05-191-6/+31
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | The following test case tried to compute the lexicographic minimum of the following set during alias analysis, which caused very long compile time: [p_0, p_1, p_2, p_3, p_4, p_5] -> { MemRef0[i0] : (517p_3 >= 70944 - 298p_2 and 256i0 >= -71199 + 298p_2 + 517p_3 and 256i0 <= -70944 + 298p_2 + 517p_3) or (409p_4 >= 57120 - 298p_2 and 256i0 >= -57375 + 298p_2 + 409p_4 and 256i0 <= -57120 + 298p_2 + 409p_4) or (104p_4 >= 17329 + 149p_2 - 50p_3 and 128i0 >= 17328 + 149p_2 - 50p_3 - 104p_4 and 128i0 <= 17455 + 149p_2 - 50p_3 - 104p_4) or (104p_4 <= 17328 + 149p_2 - 50p_3 and 128i0 >= 17201 + 149p_2 - 50p_3 - 104p_4 and 128i0 <= 17328 + 149p_2 - 50p_3 - 104p_4) or (409p_4 <= 57119 - 298p_2 and 256i0 >= -57120 + 298p_2 + 409p_4 and 256i0 <= -56865 + 298p_2 + 409p_4) or (517p_3 <= 70943 - 298p_2 and 256i0 >= -70944 + 298p_2 + 517p_3 and 256i0 <= -70689 + 298p_2 + 517p_3) or (p_1 >= 2 + 2p_0 and 298p_5 >= 70944 - 517p_3 and 256i0 >= -71199 + 517p_3 + 298p_5 and 256i0 <= -70944 + 517p_3 + 298p_5) or (p_1 >= 2 + 2p_0 and 298p_5 >= 57120 - 409p_4 and 256i0 >= -57375 + 409p_4 + 298p_5 >and 256i0 <= -57120 + 409p_4 + 298p_5) or (p_1 >= 2 + 2p_0 and 149p_5 <= -17329 >+ 50p_3 + 104p_4 and 128i0 >= 17328 - 50p_3 - 104p_4 + 149p_5 and 128i0 <= >17455 - 50p_3 - 104p_4 + 149p_5) or (p_1 >= 2 + 2p_0 and 149p_5 >= -17328 + >50p_3 + 104p_4 and 128i0 >= 17201 - 50p_3 - 104p_4 + 149p_5 and 128i0 <= 17328 >- 50p_3 - 104p_4 + 149p_5) or (p_1 >= 2 + 2p_0 and 298p_5 <= 57119 - 409p_4 and >256i0 >= -57120 + 409p_4 + 298p_5 and 256i0 <= -56865 + 409p_4 + 298p_5) or >(p_1 >= 2 + 2p_0 and 298p_5 <= 70943 - 517p_3 and 256i0 >= -70944 + 517p_3 + >298p_5 and 256i0 <= -70689 + 517p_3 + 298p_5) } We now guard the potentially expensive functions in Polly's scop analysis to gracefully bail out in case of overly long compilation times. llvm-svn: 303404
* [ScopInfo] Fix r302231 to use logical or (||). NFC.Michael Kruse2017-05-181-1/+1
| | | | | | | | | | | In r302231 we mistakenly use bitwise or (|) instead of logical or (||). This patch fixes that. Contributed-by: Sameer AbuAsal <sabuasal@codeaurora.org> Differential Revision: https://reviews.llvm.org/D33337 llvm-svn: 303386
* [IR] De-virtualize ~Value to save a vptrReid Kleckner2017-05-181-1/+1
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | Summary: Implements PR889 Removing the virtual table pointer from Value saves 1% of RSS when doing LTO of llc on Linux. The impact on time was positive, but too noisy to conclusively say that performance improved. Here is a link to the spreadsheet with the original data: https://docs.google.com/spreadsheets/d/1F4FHir0qYnV0MEp2sYYp_BuvnJgWlWPhWOwZ6LbW7W4/edit?usp=sharing This change makes it invalid to directly delete a Value, User, or Instruction pointer. Instead, such code can be rewritten to a null check and a call Value::deleteValue(). Value objects tend to have their lifetimes managed through iplist, so for the most part, this isn't a big deal. However, there are some places where LLVM deletes values, and those places had to be migrated to deleteValue. I have also created llvm::unique_value, which has a custom deleter, so it can be used in place of std::unique_ptr<Value>. I had to add the "DerivedUser" Deleter escape hatch for MemorySSA, which derives from User outside of lib/IR. Code in IR cannot include MemorySSA headers or call the MemoryAccess object destructors without introducing a circular dependency, so we need some level of indirection. Unfortunately, no class derived from User may have any virtual methods, because adding a virtual method would break User::getHungOffOperands(), which assumes that it can find the use list immediately prior to the User object. I've added a static_assert to the appropriate OperandTraits templates to help people avoid this trap. Reviewers: chandlerc, mehdi_amini, pete, dberlin, george.burgess.iv Reviewed By: chandlerc Subscribers: krytarowski, eraman, george.burgess.iv, mzolotukhin, Prazek, nlewycky, hans, inglorion, pcc, tejohnson, dberlin, llvm-commits Differential Revision: https://reviews.llvm.org/D31261 llvm-svn: 303362
* [Fortran Support] Change "global" pattern match to work for paramsSiddharth Bhat2017-05-181-44/+4
| | | | | | | | | | | | | | | | | | Summary: - Rename global / local naming convention that did not make much sense to Visible / Invisible, where the visible refers to whether the ALLOCATE call to the Fortran array is present in the current module or not. - This match now works on both cross fortran module globals and on parameters to functions since neither of them are necessarily allocated at the point of their usage. - Add testcase that matches against both a load and a store against function parameters. Differential Revision: https://reviews.llvm.org/D33190 llvm-svn: 303356
* Adjust formattingTobias Grosser2017-05-151-1/+1
| | | | llvm-svn: 303065
* [Polly][NewPM][WIP] Add a ScopPassManagerPhilip Pfaffe2017-05-152-1/+101
| | | | | | | | | | | | | | | | | | | | | | | | | | This patch adds both a ScopAnalysisManager and a ScopPassManager. The ScopAnalysisManager is itself a Function-Analysis, and manages analyses on Scops. The ScopPassManager takes care of building Scop pass pipelines. This patch is marked WIP because I've left two FIXMEs which I need to think about some more. Both of these deal with invalidation: Deferred invalidation is currently not implemented. Deferred invalidation deals with analyses which cache references to other analysis results. If these results are invalidated, invalidation needs to be propagated into the caching analyses. The ScopPassManager as implemented assumes that ScopPasses do not affect other Scops in any way. There has been some discussion about this on other patch threads, however it makes sense to reiterate this for this specific patch. I'm uploading this patch even though it's incomplete to encourage discussion and give you an impression of how this is going to work. Differential Revision: https://reviews.llvm.org/D33192 llvm-svn: 303062
* [Polly][NewPM] Port ScopInfo to the new PassManagerPhilip Pfaffe2017-05-153-20/+50
| | | | llvm-svn: 303056
* [Fortran Support] Add pattern match for Fortran Arrays that are parameters.Siddharth Bhat2017-05-152-84/+67
| | | | | | | | | | | - This breaks the previous assumption that Fortran Arrays are `GlobalValue`. - The names of functions were getting unwieldy. So, I renamed the Fortran related functions. Differential Revision: https://reviews.llvm.org/D33075 llvm-svn: 303040
* [Simplify] Fix r302986 that introduced non-inferrable templates.Siddharth Bhat2017-05-151-13/+19
| | | | | | | | | | | | | | - auto + decltype + template use was not inferrable in `Transform/Simplify.cpp accessesInOrder`. - changed code to explicitly construct required vector instead of using higher order iterator helpers. - Failing compiler spec: Apple LLVM version 7.3.0 (clang-703.0.31) Target: x86_64-apple-darwin15.6.0 llvm-svn: 303039
OpenPOWER on IntegriCloud