summaryrefslogtreecommitdiffstats
path: root/clang/test/OpenMP
Commit message (Collapse)AuthorAgeFilesLines
...
* [OPENMP] Fix assert fail after target implicit map checks.Alexey Bataev2017-12-051-0/+4
| | | | | | | | If the error is generated during analysis of implicitly or explicitly mapped variables, it may cause compiler crash because of incorrect analysis. llvm-svn: 319774
* [OPENMP] Fix PR35486: crash when collapsing loops with dependent iteration ↵Alexey Bataev2017-12-0427-16/+94
| | | | | | | | | | spaces. Though it is incorrect from point of view of OpenMP standard to have dependent iteration space in OpenMP loops, compiler should not crash. Patch fixes this problem. llvm-svn: 319700
* [OpenMP] Initial implementation of code generation for pragma 'teams ↵Carlo Bertolli2017-12-0412-1/+3259
| | | | | | | | | | distribute parallel for simd' on host https://reviews.llvm.org/D40795 This includes regression tests for all associated clauses. llvm-svn: 319696
* [OPENMP] Codegen for `distribute simd` directive.Alexey Bataev2017-12-046-0/+1461
| | | | | | Initial codegen support for `distribute simd` directive. llvm-svn: 319661
* [OPENMP] Emit `__tgt_target_teams` for all teams directives.Alexey Bataev2017-12-0119-93/+93
| | | | | | | | Previously we emitted `__tgt_target_teams` only for standalone teams directives. This patch allows emit this function for all teams-based directives. llvm-svn: 319585
* [OPENMP] Do not allow variables to be first|last-privates inAlexey Bataev2017-12-0139-92/+155
| | | | | | | | distribute directives. OpenMP standard does not allow to mark the variables as firstprivate and lastprivate at the same time in distribute-based directives. Patch fixes this problem. llvm-svn: 319560
* [OpenMP] Diagnose undeclared variables on declare target clauseKelvin Li2017-11-301-0/+4
| | | | | | | | | | | | | Clang asserts on undeclared variables on the to or link clause in the declare target directive. The patch is to properly diagnose the error. // foo1 and foo2 are not declared #pragma omp declare target to(foo1) #pragma omp declare target link(foo2) Differential Revision: https://reviews.llvm.org/D40588 llvm-svn: 319458
* [OPENMP] Fix possible assert for target regions with incorrect innerAlexey Bataev2017-11-305-0/+15
| | | | | | | | | teams region. If the inner teams region is not correct, it may cause an assertion when processing outer target region. Patch fixes this problem. llvm-svn: 319450
* [OPENMP] Allow only loop control variables in distribute simdAlexey Bataev2017-11-2914-440/+174
| | | | | | | | | directives. According to the OpenMP standard, only loop control variables can be used in linear clauses of distribute-based simd directives. llvm-svn: 319362
* [OPENMP] Do not allow `linear` clauses on non-simd distributeAlexey Bataev2017-11-298-890/+11
| | | | | | | | directives. `linear` clause is not allowed on non-simd distribute-based directives. llvm-svn: 319332
* [OPENMP] Generalize capturing of clauses expressions.Alexey Bataev2017-11-281-7/+3
| | | | | | | The handling and capturing of the non-constant expressions of some of the capturable clauses in combined directives is generalized. llvm-svn: 319227
* [OPENMP] Codegen for `distribute parallel for simd` directive.Alexey Bataev2017-11-277-0/+4240
| | | | | | | Initial codegen for `#pragma omp distribute parallel for simd` directive and its clauses. llvm-svn: 319079
* [OPENMP] Improve handling of cancel directives in target-basedAlexey Bataev2017-11-271-0/+2
| | | | | | | | | constructs, NFC. Improved handling of cancel|cancellation point directives inside target-based for directives. llvm-svn: 319046
* [OPENMP] Add support for cancel constructs in `target teams distributeAlexey Bataev2017-11-221-1/+4
| | | | | | | | | parallel for`. Add support for cancel/cancellation point directives inside `target teams distribute parallel for` directives. llvm-svn: 318881
* [OPENMP] Add support for cancel constructs in [teams] distributeAlexey Bataev2017-11-224-0/+22
| | | | | | | | | parallel for directives. Added codegen/sema support for cancel constructs in [teams] distribute parallel for directives. llvm-svn: 318872
* [OPENMP] Added missed checks for for [simd] based directives.Alexey Bataev2017-11-225-2/+352
| | | | | | | Added missed checks/analysis for safelen/simdlen clauses + linear clause in for [simd] based directives. llvm-svn: 318860
* [OPENMP] Do not mark captured variables as artificial in debug info.Alexey Bataev2017-11-221-0/+10
| | | | | | | Captured variables should not be marked as artificial parameters in outlined functions in debug info. llvm-svn: 318843
* [OpenMP] Adjust arguments of nvptx runtime functionsJonas Hahnfeld2017-11-224-13/+13
| | | | | | | | | | | | | In the future the compiler will analyze whether the OpenMP runtime needs to be (fully) initialized and avoid that overhead if possible. The functions already take an argument to transfer that information to the runtime, so pass in the default value 1. (This is needed for binary compatibility with libomptarget-nvptx currently being upstreamed.) Differential Revision: https://reviews.llvm.org/D40354 llvm-svn: 318836
* [OPENMP] Codegen for `target teams` directive.Alexey Bataev2017-11-222-9/+36
| | | | | | Added codegen of the clauses for `target teams` directive. llvm-svn: 318834
* [OpenMP] Fix tests after r318789Richard Trieu2017-11-214-15/+15
| | | | | | Update use of __tgt_target that had some 32bit types updated to 64bit. llvm-svn: 318811
* [Clang][OpenMP] New clang/libomptarget map interface: new function ↵George Rokos2017-11-2130-564/+576
| | | | | | | | | | | signatures, clang-side This clang patch changes the __tgt_* API function signatures in preparation for the new map interface. Changes are: Device IDs 32bits --> 64bits, Flags 32bits --> 64bits Differential revision: https://reviews.llvm.org/D40281 llvm-svn: 318789
* Fix test/OpenMP/nvptx_data_sharing.cppJonas Hahnfeld2017-11-211-2/+2
| | | | | | This was an oversight that stayed in the test from development. llvm-svn: 318779
* [OpenMP] Add implicit data sharing support when offloading to NVIDIA GPUs ↵Gheorghe-Teodor Bercea2017-11-213-13/+65
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | using OpenMP device offloading Summary: This patch is part of the development effort to add support in the current OpenMP GPU offloading implementation for implicitly sharing variables between a target region executed by the team master thread and the worker threads within that team. This patch is the first of three required for successfully performing the implicit sharing of master thread variables with the worker threads within a team. The remaining two patches are: - Patch D38978 to the LLVM NVPTX backend which ensures the lowering of shared variables to an device memory which allows the sharing of references; - Patch (coming soon) is a patch to libomptarget runtime library which ensures that a list of references to shared variables is properly maintained. A simple code snippet which illustrates an implicit data sharing situation is as follows: ``` #pragma omp target { // master thread only int v; #pragma omp parallel { // worker threads // use v } } ``` Variable v is implicitly shared from the team master thread which executes the code in between the target and parallel directives. The worker threads must operate on the latest version of v, including any updates performed by the master. The code generated in this patch relies on the LLVM NVPTX patch (mentioned above) which prevents v from being lowered in the thread local memory of the master thread thus making the reference to this variable un-shareable with the workers. This ensures that the code generated by this patch is correct. Since the parallel region is outlined the passing of arguments to the outlined regions must preserve the original order of arguments. The runtime therefore maintains a list of references to shared variables thus ensuring their passing in the correct order. The passing of arguments to the outlined parallel function is performed in a separate function which the data sharing infrastructure constructs in this patch. The function is inlined when optimizations are enabled. Reviewers: hfinkel, carlo.bertolli, arpith-jacob, Hahnfeld, ABataev, caomhin Reviewed By: ABataev Subscribers: cfe-commits, jholewinski Differential Revision: https://reviews.llvm.org/D38976 llvm-svn: 318773
* [OpenMP] Initial implementation of code generation for pragma 'teams ↵Carlo Bertolli2017-11-2012-0/+3389
| | | | | | | | | | distribute parallel for' on host https://reviews.llvm.org/D40187 This patch implements code gen for 'teams distribute parallel for' on the host, including all its clauses and related regression tests. llvm-svn: 318692
* [OpenMP] Show error if VLAs are not supportedJonas Hahnfeld2017-11-181-0/+201
| | | | | | | | | | | | | | Some target devices (e.g. Nvidia GPUs) don't support dynamic stack allocation and hence no VLAs. Print errors with description instead of failing in the backend or generating code that doesn't work. This patch handles explicit uses of VLAs (local variable in target or declare target region) or implicitly generated (private) VLAs for reductions on VLAs or on array sections with non-constant size. Differential Revision: https://reviews.llvm.org/D39505 llvm-svn: 318601
* [OPENMP] Codegen for `target simd` construct.Alexey Bataev2017-11-173-0/+1194
| | | | | | Added codegen support for `target simd` directive. llvm-svn: 318536
* [OPENMP] Add support for cancelling inside target parallel forAlexey Bataev2017-11-161-1/+3
| | | | | | | | directive. Added missed support for cancelling of target parallel for construct. llvm-svn: 318434
* [OPENMP] Fix DSA analysis for threadprivates after deserialization.Alexey Bataev2017-11-141-4/+7
| | | | | | | If threadprivate vaible is deserialized, it is not marked as threadprivate in DSAStack. llvm-svn: 318194
* [NFC] Pacify bbot for OpenMP 'teams distribute parallel for'Carlo Bertolli2017-11-101-1/+1
| | | | llvm-svn: 317898
* [OpenMP] Parse+Sema for copyin clause of 'teams distribute parallel for'Carlo Bertolli2017-11-102-8/+124
| | | | | | | | https://reviews.llvm.org/D39902 Simply leverage existing implementation and verify correct functioning with two regression tests. llvm-svn: 317893
* [OpenMP] diagnose assign to firstprivate const, patch by Joel E. DennyAlexey Bataev2017-11-101-1/+3
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | Summary: [OpenMP] diagnose assign to firstprivate const Clang does not diagnose assignments to const variables declared firstprivate. Furthermore, codegen is broken such that, at run time, such assignments simply have no effect. For example, the following prints 0 not 1: int main() { const int i = 0; #pragma omp parallel firstprivate(i) { i=1; printf("%d\n", i); } return 0; } This commit makes these assignments a compile error, which is consistent with other OpenMP compilers I've tried (pgcc 17.4-0, gcc 6.3.0). Reviewers: ABataev Reviewed By: ABataev Subscribers: cfe-commits Differential Revision: https://reviews.llvm.org/D39859 llvm-svn: 317891
* [OPENMP] Codegen for `#pragma omp target parallel for simd`.Alexey Bataev2017-11-093-186/+1285
| | | | | | Added codegen for `#pragma omp target parallel for simd` and clauses. llvm-svn: 317813
* [OPENMP] Treat '#pragma omp target parallel for simd' as simd directive.Alexey Bataev2017-11-096-43/+49
| | | | | | | `#pragma omp target parallel for simd` mistakenly was not treated as a simd directive, fixed this problem. llvm-svn: 317811
* [OPENMP] Codegen for `#pragma omp target parallel for`.Alexey Bataev2017-11-084-186/+1394
| | | | llvm-svn: 317719
* [OPENMP] Fix PR35152: Do not use getInvokeDest() function for EH checks.Alexey Bataev2017-11-021-6/+29
| | | | | | | The compiler may crash under some conditions if the getInvokeDest() is used, but later it is not used. Fixed this problem in OpenMP. llvm-svn: 317227
* [OPENMP] Fix PR35156: Get correct thread id with windows exceptions.Alexey Bataev2017-11-0211-10/+48
| | | | | | | If the thread id is requested in windows mode within funclets, we may generate incorrect function call that could lead to broken codegen. llvm-svn: 317208
* [OpenMP] Extend "Avoid VLAs for reduction" optimization to VLAs as baseJonas Hahnfeld2017-11-021-12/+46
| | | | | | | | | We can generate constant sized arrays whenever the array section has constant length, even if the base expression itself is a VLA. Differential Revision: https://reviews.llvm.org/D39504 llvm-svn: 317207
* [OPENMP] Improve debug info for taskgroup implicitly generatedAlexey Bataev2017-10-251-9/+11
| | | | | | expressions. llvm-svn: 316585
* [OPENMP] Fix PR35013: Fix passing VLAs captures to outlined functions.Alexey Bataev2017-10-247-13/+33
| | | | | | | Fixed passing of VLAs and variably-modified types to outlined functions. Synchronized passing with the types codegen. llvm-svn: 316488
* [OpenMP] Avoid VLAs for some reductions on array sectionsJonas Hahnfeld2017-10-232-100/+322
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | In some cases the compiler can deduce the length of an array section as constants. With this information, VLAs can be avoided in place of a constant sized array or even a scalar value if the length is 1. Example: int a[4], b[2]; pragma omp parallel reduction(+: a[1:2], b[1:1]) { } For chained array sections, this optimization is restricted to cases where all array sections except the last have a constant length 1. This trivially guarantees that there are no holes in the memory region that needs to be privatized. Example: int c[3][4]; pragma omp parallel reduction(+: c[1:1][1:2]) { } This relands commit r316229 that I reverted in r316235 because it failed on some bots. During investigation I found that this was because Clang and GCC evaluate the two arguments to emplace_back() in ReductionCodeGen::emitSharedLValue() in a different order, hence leading to a different order of generated instructions in the final LLVM IR. Fix this by passing in the arguments from temporary variables that are evaluated in a defined order. Differential Revision: https://reviews.llvm.org/D39136 llvm-svn: 316362
* Revert "[OpenMP] Avoid VLAs for some reductions on array sections"Jonas Hahnfeld2017-10-202-322/+100
| | | | | | | | | | This breaks at least two buildbots: http://lab.llvm.org:8011/builders/clang-cmake-x86_64-avx2-linux/builds/1175 http://lab.llvm.org:8011/builders/clang-atom-d525-fedora-rel/builds/10478 This reverts commit r316229 during local investigation. llvm-svn: 316235
* [OpenMP] Avoid VLAs for some reductions on array sectionsJonas Hahnfeld2017-10-202-100/+322
| | | | | | | | | | | | | | | | | | | | | | | In some cases the compiler can deduce the length of an array section as constants. With this information, VLAs can be avoided in place of a constant sized array or even a scalar value if the length is 1. Example: int a[4], b[2]; pragma omp parallel reduction(+: a[1:2], b[1:1]) { } For chained array sections, this optimization is restricted to cases where all array sections except the last have a constant length 1. This trivially guarantees that there are no holes in the memory region that needs to be privatized. Example: int c[3][4]; pragma omp parallel reduction(+: c[1:1][1:2]) { } Differential Revision: https://reviews.llvm.org/D39136 llvm-svn: 316229
* [OPENMP] Fix capturing of boolean variables in debug mode.Alexey Bataev2017-10-171-2/+2
| | | | | | | | If the variables is boolean and we generating inner function with real types, the codegen may crash because of not loading boolean value from memory. llvm-svn: 316011
* [OpenMP] Implement omp_is_initial_device() as builtinJonas Hahnfeld2017-10-171-0/+36
| | | | | | | | This allows to return the static value that we know at compile time. Differential Revision: https://reviews.llvm.org/D38968 llvm-svn: 316001
* [OPENMP] Fix PR34927: Emit initializer for reduction array with declareAlexey Bataev2017-10-121-0/+20
| | | | | | | | | reduction. If the reduction is an array or an array section and reduction operation is declare reduction without initializer, it may lead to crash. llvm-svn: 315611
* [OPENMP] Fix PR34926: Fix handling of the array sections passed asAlexey Bataev2017-10-121-0/+8
| | | | | | | | | function params. Codegen could crash if the array section base expression is the function parameter. llvm-svn: 315586
* [OPENMP] Fix PR34925: Fix getting thread_id lvalue for inlined regionsAlexey Bataev2017-10-121-0/+30
| | | | | | | | | | | in C. If we try to get the lvalue for thread_id variables in inlined regions, we did not use the correct version of function. Fixed this bug by adding overrided version of the function getThreadIDVariableLValue for inlined regions. llvm-svn: 315578
* [OPENMP] Fix PR34916: Crash on mixing taskloop|tasks directives.Alexey Bataev2017-10-111-0/+4
| | | | | | | | If both taskloop and task directives are used at the same time in one program, we may ran into the situation when the particular type for task directive is reused for taskloop directives. Patch fixes this problem. llvm-svn: 315464
* [OPENMP] Add default codegen|tests for 'target parallel for[ simd]'Alexey Bataev2017-10-102-0/+464
| | | | | | | | | constructs. Added default codegen for 'target parallel for' construct + tests for default codegen of 'target parallel for[ simd]' constructs. llvm-svn: 315315
* [OPENMP] Do not capture local static variables.Alexey Bataev2017-10-061-1/+38
| | | | | | | | Previously we may erroneously try to capture locally declared static variables, which will lead to crash for target-based constructs. Patch fixes this problem. llvm-svn: 315076
OpenPOWER on IntegriCloud