summaryrefslogtreecommitdiffstats
path: root/clang/test/OpenMP
Commit message (Collapse)AuthorAgeFilesLines
...
* Revert "[OPENMP]Initial support for the delayed diagnostics."Alexey Bataev2019-02-081-13/+0
| | | | | | | This reverts commit r353540. Erroneously committed, need to fix the message and description. llvm-svn: 353541
* [OPENMP]Initial support for the delayed diagnostics.Alexey Bataev2019-02-081-0/+13
| | | | | | | | | It is important to delay the emission of the diagnostic messages for the functions unless it is proved that the function is going to be used on the device side. It is required to support compilation with some of the target-specific system headers. llvm-svn: 353540
* [opaque pointer types] Cleanup CGBuilder's Create*GEP.James Y Knight2019-02-0815-58/+57
| | | | | | | | | | | | | | | | | | | | | | | Some of these functions take some extraneous arguments, e.g. EltSize, Offset, which are computable from the Type and DataLayout. Add some asserts to ensure that the computed values are consistent with the passed-in values, in preparation for eliminating the extraneous arguments. This also asserts that the Type is an Array for the calls named "Array" and a Struct for the calls named "Struct". Then, correct a couple of errors: 1. Using CreateStructGEP on an array type. (this causes the majority of the test differences, as struct GEPs are created with i32 indices, while array GEPs are created with i64 indices) 2. Passing the wrong Offset to CreateStructGEP in TargetInfo.cpp on x86-64 NACL (which uses 32-bit pointers). Differential Revision: https://reviews.llvm.org/D57766 llvm-svn: 353529
* [OPENMP] issue error messages for multiple teams contructs in a target constructKelvin Li2019-02-051-0/+14
| | | | | | | | | | | | | | | | The fix is to issue error messages if there are more than one teams construct inside a target constructs. #pragma omp target { #pragma omp teams { ... } #pragma omp teams { ... } } llvm-svn: 353186
* [OpenMP] Adding support to the mutexinoutset dep-typeSergi Mateo Bellido2019-02-0415-68/+107
| | | | | | | | | | | Summary: this commit adds support to a new dependence type introduced in OpenMP 5.0. The LLVM OpenMP RTL already supports this feature, so we only need to modify CLANG to take advantage of them. Differential Revision: https://reviews.llvm.org/D57576 llvm-svn: 353018
* [OpenMP 5.0] Parsing/sema support for "omp declare mapper" directive.Michael Kruse2019-02-014-0/+256
| | | | | | | | | | | | | | | | | This patch implements parsing and sema for "omp declare mapper" directive. User defined mapper, i.e., declare mapper directive, is a new feature in OpenMP 5.0. It is introduced to extend existing map clauses for the purpose of simplifying the copy of complex data structures between host and device (i.e., deep copy). An example is shown below: struct S { int len; int *d; }; #pragma omp declare mapper(struct S s) map(s, s.d[0:s.len]) // Memory region that d points to is also mapped using this mapper. Contributed-by: Lingda Li <lildmh@gmail.com> Differential Revision: https://reviews.llvm.org/D56326 llvm-svn: 352906
* [OPENMP]Fix PR40536: Do not emit __kmpc_push_target_tripcount if notAlexey Bataev2019-01-303-2/+5
| | | | | | | | | | required. Function __kmpc_push_target_tripcount should be emitted only if the offloading entry is going to be emitted (for use in tgt_target... functions). Otherwise, it should not be emitted. llvm-svn: 352669
* [OPENMP]Fix PR40513: lastprivate taskloop counter.Alexey Bataev2019-01-291-2/+15
| | | | | | | | We don't need to use the predetermined data-sharing attributes for the loop counters if the user explicitly specified correct data-sharing attributes for such variables. llvm-svn: 352543
* [OPENMP]Make the loop with unsigned counter countable.Alexey Bataev2019-01-293-5/+10
| | | | | | | | According to the report, better to keep the original strict compare operation as the loop condition with unsigned loop counters to make the loop countable. This allows further loop transformations. llvm-svn: 352526
* [OPENMP][NVPTX]Emit service debug variable for NVPTX.Alexey Bataev2019-01-281-0/+1
| | | | | | | In case of the empty module, the ptxas tool may emit error message about empty debug info sections. This patch fixes this bug. llvm-svn: 352421
* Emit !callback metadata and introduce the callback attributeJohannes Doerfert2019-01-191-3/+4
| | | | | | | | | | | | | | | | | | | | | | | | | | | With commit r351627, LLVM gained the ability to apply (existing) IPO optimizations on indirections through callbacks, or transitive calls. The general idea is that we use an abstraction to hide the middle man and represent the callback call in the context of the initial caller. It is described in more detail in the commit message of the LLVM patch r351627, the llvm::AbstractCallSite class description, and the language reference section on callback-metadata. This commit enables clang to emit !callback metadata that is understood by LLVM. It does so in three different cases: 1) For known broker functions declarations that are directly generated, e.g., __kmpc_fork_call for the OpenMP pragma parallel. 2) For known broker functions that are identified by their name and source location through the builtin detection, e.g., pthread_create from the POSIX thread API. 3) For user annotated functions that carry the "callback(callee, ...)" attribute. The attribute has to include the name, or index, of the callback callee and how the passed arguments can be identified (as many as the callback callee has). See the callback attribute documentation for detailed information. Differential Revision: https://reviews.llvm.org/D55483 llvm-svn: 351629
* [OpenMP] Avoid remainder operations for loop index values on a collapsed ↵Gheorghe-Teodor Bercea2019-01-094-36/+188
| | | | | | | | | | | | | | | | loop nest. Summary: Change the strategy for computing loop index variables after collapsing a loop nest via the collapse clause by replacing the expensive remainder operation with multiplications and additions. Reviewers: ABataev, caomhin Reviewed By: ABataev Subscribers: guansong, arphaman, cfe-commits Differential Revision: https://reviews.llvm.org/D56413 llvm-svn: 350759
* [OpenMP] Add flag for preventing the extension to 64 bits for the collapse ↵Gheorghe-Teodor Bercea2019-01-091-10/+15
| | | | | | | | | | | | | | | | loop counter Summary: Introduce a compiler flag for cases when the user knows that the collapsed loop counter can be safely represented using at most 32 bits. This will prevent the emission of expensive mathematical operations (such as the div operation) on the iteration variable using 64 bits where 32 bit operations are sufficient. Reviewers: ABataev, caomhin Reviewed By: ABataev Subscribers: hfinkel, kkwli0, guansong, cfe-commits Differential Revision: https://reviews.llvm.org/D55928 llvm-svn: 350758
* Incorrect implicit data-sharing for nested tasksAlexey Bataev2019-01-091-1/+19
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | Summary: There is a minor issue in how the implicit data-sharings for nested tasks are computed. For the following example: ``` int x; #pragma omp task shared(x) #pragma omp task x++; ``` We compute an implicit data-sharing of shared for `x` in the second task although I think that it should be firstprivate. Below you can find the part of the OpenMP spec that covers this example: - // In a task generating construct, if no default clause is present, a variable for which the data-sharing attribute is not determined by the rules above and that in the enclosing context is determined to be shared by all implicit tasks bound to the current team is shared.// - //In a task generating construct, if no default clause is present, a variable for which the data-sharing attribute is not determined by the rules above is firstprivate.// Since each implicit-task has its own copy of `x`, we shouldn't apply the first rule. Reviewers: ABataev Reviewed By: ABataev Subscribers: cfe-commits, rogfer01 Tags: #openmp Differential Revision: https://reviews.llvm.org/D56430 llvm-svn: 350734
* [OPENMP]Fix PR40191: Do not allow orphaned cancellation constructs.Alexey Bataev2019-01-082-40/+40
| | | | | | Prohibited use of the orphaned cancellation directives. llvm-svn: 350634
* [OPENMP]Add call to __kmpc_push_target_tripcount() function.Alexey Bataev2019-01-078-2/+37
| | | | | | | | | | Each we create the target regions with the teams distribute inner region, we can better estimate number of the teams required to execute the target region. Function __kmpc_push_target_tripcount() is used for purpose, which accepts device_id and the number of the iterations, performed by the associated loop. llvm-svn: 350571
* [OPENMP][NVPTX]Reduce number of barriers in reductions.Alexey Bataev2019-01-072-5/+0
| | | | | | | After the fix for the syncthreads we don't need to generate extra barriers for the parallel reductions. llvm-svn: 350530
* [OpenMP] Refactor const restriction for linearJoel E. Denny2019-01-0415-30/+30
| | | | | | | | | | | | | | As discussed in D56113, this patch refactors the implementation of the const restriction for linear to reuse a function introduced by D56113. A side effect is that, if a variable has mutable members, this diagnostic is now skipped, and the diagnostic for the variable not being an integer or pointer is reported instead. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D56299 llvm-svn: 350441
* [OpenMP] Refactor const restriction for reductionsJoel E. Denny2019-01-0432-448/+448
| | | | | | | | | | | | | As discussed in D56113, this patch refactors the implementation of the const restriction for reductions to reuse a function introduced by D56113. A side effect is that diagnostics sometimes now say "variable" instead of "list item" when a list item is a variable. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D56298 llvm-svn: 350440
* [OpenMP] Replace predetermined shared for const variableJoel E. Denny2019-01-0438-349/+357
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | The following appears in OpenMP 3.1 sec. 2.9.1.1 as a predetermined data-sharing attribute: > Variables with const-qualified type having no mutable member are > shared. It does not appear in OpenmP 4.0, 4.5, or 5.0. This patch removes the implementation of that attribute when the requested OpenMP version is greater than 3.1. One effect of that removal is that `default(none)` affects const variables without mutable members. Also, without this patch, if a const variable without mutable members was explicitly lastprivate or private, it was an error because it was predetermined shared. Now, clang instead complains that it's const without mutable fields, which is a more intelligible diagnostic. That should be fine for all of the above versions because they all have something like the following, which is quoted from OpenMP 5.0 sec. 2.19.3: > A variable that is privatized must not have a const-qualified type > unless it is of class type with a mutable member. This restriction does > not apply to the firstprivate clause. reduction and linear clauses already have separate checks for const variables. Future patches will merge the implementations. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D56113 llvm-svn: 350439
* [OPENMP][NVPTX]Use new functions from the runtime library.Alexey Bataev2019-01-0411-24/+24
| | | | | | Updated codegen to use the new functions from the runtime library. llvm-svn: 350415
* [OPENMP][NVPTX]Use __kmpc_barrier_simple_spmd(nullptr, 0) instead ofAlexey Bataev2019-01-036-51/+51
| | | | | | | | | | nvvm_barrier0. Use runtime functions instead of the direct call to the nvvm intrinsics. It allows to prevent some dangerous LLVM optimizations, that breaks the code for the NVPTX target. llvm-svn: 350328
* [OpenMP] Added support for explicit mapping of classes using 'this' pointer. ↵Patrick Lyster2019-01-023-2/+95
| | | | | | Differential revision: https://reviews.llvm.org/D55982 llvm-svn: 350252
* [OPENMP]Fix processing of the clauses on target combined directives.Alexey Bataev2018-12-281-0/+21
| | | | | | | | For constants with the predefined data-sharing clauses we may had troubles with the target combined directives. It may cause compiler crash in some corner cases. llvm-svn: 350127
* Pass a concrete triple for two OpenMP tests that depend on TLSNico Weber2018-12-262-26/+26
| | | | | | | | | | | | | | | Not all %itanium_abi_triple values support TLS. Makes OpenMP/declare_reduction_codegen.cpp, OpenMP/parallel_copyin_codegen.cpp for %itanium_abi_triples without TLS support. Alternatively we could pass -fnoopenmp-use-tls and tweak some of the CHECK lines, but this seems a bit simpler. Fixes PR40156. Differential Revision: https://reviews.llvm.org/D56086 llvm-svn: 350067
* [CodeGen] Generate llvm.loop.parallel_accesses instead of ↵Michael Kruse2018-12-2011-221/+228
| | | | | | | | | | | | | | llvm.mem.parallel_loop_access metadata. Instead of generating llvm.mem.parallel_loop_access metadata, generate llvm.access.group on instructions and llvm.loop.parallel_accesses on loops. There is one access group per generated loop. This is clang part of D52116/r349725. Differential Revision: https://reviews.llvm.org/D52117 llvm-svn: 349823
* [OPENMP]Mark the loop as started when initialized.Alexey Bataev2018-12-191-1/+12
| | | | | | | | | Need to mark the loop as started when the initialization statement is found. It is required to prevent possible incorrect loop iteraton variable detection during template instantiation and fix the compiler crash during the codegen. llvm-svn: 349657
* [OpenMP] Fix data sharing analysis in nested clauseJoel E. Denny2018-12-191-0/+10
| | | | | | | | | | | | | | | | | | | | | | Without this patch, clang doesn't complain that X needs explicit data sharing attributes in the following: ``` #pragma omp target teams default(none) { #pragma omp parallel num_threads(X) ; } ``` However, clang does produce that complaint after the braces are removed. With this patch, clang complains in both cases. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D55861 llvm-svn: 349635
* [OPENMP] parsing and sema support for 'close' map-type-modifierKelvin Li2018-12-1812-24/+170
| | | | | | | | | | | | A map clause with the close map-type-modifier is a hint to prefer that the variables are mapped using a copy into faster memory. Patch by Ahsan Saghir (saghir) Differential Revision: https://reviews.llvm.org/D55719 llvm-svn: 349551
* [OPENMP][NVPTX]Emit shared memory buffer for reduction as 128 bytesAlexey Bataev2018-12-1810-11/+11
| | | | | | | | | | | buffer. Seems to me, nvlink has a bug with the proper support of the weakly linked symbols. It does not allow to define several shared memory buffer with the different sizes even with the weak linkage. Instead we always use 128 bytes buffer to prevent nvlink from the error message emission. llvm-svn: 349540
* [OPENMP][NVPTX]Added extra sync point to the inter-warp copy function.Alexey Bataev2018-12-181-0/+4
| | | | | | | The parallel reduction operation requires an extra synchronization point in the inter-warp copy function to avoid divergence. llvm-svn: 349525
* [OPENMP][NVPTX]Improved interwarp copy function.Alexey Bataev2018-12-142-28/+21
| | | | | | | | | Inlined runtime with the current implementation of the interwarp copy function leads to the undefined behavior because of the not quite correct implementation of the barriers. Start using generic __kmpc_barier function instead of the custom made barriers. llvm-svn: 349192
* [OPENMP][NVPTX] Fix globalization of the mapped array sections.Alexey Bataev2018-12-061-19/+52
| | | | | | | | | If the array section is based on pointer and this sections is mapped in target region + then it is used in the inner parallel region, it also must be globalized as the pointer itself is passed by value, not by reference. llvm-svn: 348492
* [OPENMP][NVPTX]Fixed emission of the critical region.Alexey Bataev2018-12-041-1/+7
| | | | | | | | | | | | | Critical regions in NVPTX are the constructs, which, generally speaking, are not supported by the NVPTX target. Instead we're using special technique to handle the critical regions. Currently they are supported only within the loop and all the threads in the loop must execute the same critical region. Inside of this special regions the regions still must be emitted as critical, to avoid possible data races between the teams + synchronization must use __kmpc_barrier functions. llvm-svn: 348272
* [OPENMP][NVPTX]Mark __kmpc_barrier functions as convergent.Alexey Bataev2018-12-041-2/+8
| | | | | | | | __kmpc_barrier runtime functions must be marked as convergent to prevent some dangerous optimizations. Also, for NVPTX target all barriers must be emitted as simple barriers. llvm-svn: 348271
* Move AST tests into their own test directory; NFC.Aaron Ballman2018-11-301-84/+0
| | | | | | This moves everything primarily testing the functionality of -ast-dump and -ast-print into their own directory, rather than leaving the tests spread around the testing directory. llvm-svn: 348017
* [OPENMP][NVPTX]Call get __kmpc_global_thread_num in worker afterAlexey Bataev2018-11-291-1/+1
| | | | | | | | | initialization. Function __kmpc_global_thread_num should be called only after initialization, not earlier. llvm-svn: 347919
* [OpenMP] Add a new version of the SPMD deinit kernel functionGheorghe-Teodor Bercea2018-11-298-20/+20
| | | | | | | | | | | | | | Summary: This patch adds a new runtime for the SPMD deinit kernel function which replaces the previous function. The new function takes as argument the flag which signals whether the runtime is required or not. This enables the compiler to optimize out the part of the deinit function which are not needed. Reviewers: ABataev, caomhin Reviewed By: ABataev Subscribers: jholewinski, guansong, cfe-commits Differential Revision: https://reviews.llvm.org/D54970 llvm-svn: 347915
* [OPENMP]Fix emission of the target regions in virtual functions.Alexey Bataev2018-11-281-0/+24
| | | | | | | | Fixed emission of the target regions found in the virtual functions. Previously we may end up with the situation when those regions could be skipped. llvm-svn: 347793
* [OPENMP][NVPTX]Basic support for reductions across the teams.Alexey Bataev2018-11-271-948/+7
| | | | | | Added basic codegen support for the reductions across the teams. llvm-svn: 347715
* [OPENMP][NVPTX]Emit default locations with the correct Exec|RuntimeAlexey Bataev2018-11-263-29/+173
| | | | | | | | | | | modes. If the region is inside target|teams|distribute region, we can emit the locations with the correct info for execution mode and runtime mode. Patch adds this ability to the NVPTX codegen to help the optimizer to produce better code. llvm-svn: 347583
* [OPENMP][NVPTX]Emit default locations as constant with undefined mode.Alexey Bataev2018-11-212-1/+8
| | | | | | | | | | | | For the NVPTX target default locations should be emitted as constants + additional info must be emitted in the reserved_2 field of the ident_t structure. The 1st bit controls the execution mode and the 2nd bit controls use of the lightweight runtime. The combination of the bits for Non-SPMD mode + lightweight runtime represents special undefined mode, used outside of the target regions for orphaned directives or functions. Should allow and additional optimization inside of the target regions. llvm-svn: 347425
* [OPENMP]Fix handling of the LCVs in loop-based directives.Alexey Bataev2018-11-211-0/+90
| | | | | | | | | Loop-control variables with the default data-sharing attributes should not be captured in the OpenMP region as they are private by default. Also, default attributes should be emitted for such variables in the inner OpenMP regions for the correct data sharing during codegen. llvm-svn: 347409
* [OPENMP] Support relational-op != (not-equal) as one of the canonical Kelvin Li2018-11-2122-20/+123
| | | | | | | | | | | | | | forms of random access iterator In OpenMP 4.5, only 4 relational operators are supported: <, <=, >, and >=. This work is to enable support for relational operator != (not-equal) as one of the canonical forms. Patch by Anh Tuyen Tran Differential Revision: https://reviews.llvm.org/D54441 llvm-svn: 347405
* [OpenMP] Update CHECK-DAG usage in target_parallel_codegen.cppJoel E. Denny2018-11-201-37/+33
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. There were two issues in this test: 1. There were sets of patterns for store instructions in which a pattern X could match a superset of a pattern Y. While X appeared before Y, Y's intended match appeared before X's intended match. The result was that X matched Y's intended match. Under the old overlapping behavior, Y also matched Y's intended match. Under the new non-overlapping behavior, Y had nothing left to match. This patch fixes this by gathering these sets in one place and putting the most specific patterns (Y) before the more general patterns (X). 2. The CHECK-DAG patterns involving the variables CBPADDR3 and CBPADDR4 were the same, but there was only one match in the text, so CBPADDR4 patterns had nothing to match under the new non-overlapping behavior. Moreover, a preceding related series of directives had variables (SADDR0, BPADDR0, etc.) numbered only 0 through 4, but this series had variables numbered 0 through 5. Assuming CBPADDR4's directives were not intended, this patch removes them. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54765 llvm-svn: 347351
* [OpenMP] Update CHECK-DAG usage in for_codegen.cppJoel E. Denny2018-11-201-14/+12
| | | | | | | | | | | | | | | | | | | | | This patch adjusts a test not to depend on deprecated FileCheck behavior that permits overlapping matches within a block of CHECK-DAG directives. Thus, this patch also removes uses of FileCheck's -allow-deprecated-dag-overlap command-line option. Specifically, the FileCheck variables DBG_LOC_START, DBG_LOC_END, and DBG_LOC_CANCEL were all set to the same value. As a result, three TERM_DEBUG-DAG patterns, one for each variable, all matched the same text under the old overlapping behavior. Under the new non-overlapping behavior, that's not permitted. This patch's solution is to replace these variables with one variable and replace these patterns with one pattern. Reviewed By: ABataev Differential Revision: https://reviews.llvm.org/D54764 llvm-svn: 347350
* [OpenMP] Check target architecture supports unified shared memory for ↵Patrick Lyster2018-11-191-0/+25
| | | | | | requires directive. Differential Review: https://reviews.llvm.org/D54493 llvm-svn: 347214
* [OPENMP]Fix PR39694: do not capture `this` in non-`this` region.Alexey Bataev2018-11-161-9/+18
| | | | | | | | | If lambda is used inside of the OpenMP region and captures `this`, we should recapture it in the OpenMP region also. But we should do this only if the OpenMP region is used in the context of the same class, just like the lambda. llvm-svn: 347096
* [OPENMP][NVPTX]Emit correct reduction code for teams/parallelAlexey Bataev2018-11-169-119/+414
| | | | | | | | | | | reductions. Fixed previously committed code for the reduction support in teams/parallel constructs taking into account new design of the NVPTX support in the compiler. Teams reduction are not fully functional yet, it is going to be fixed in the following patches. llvm-svn: 347081
* [OPENMP][NVPTX]Extend number of constructs executed in SPMD mode.Alexey Bataev2018-11-097-12/+23
| | | | | | | | | If the statements between target|teams|distribute directives does not require execution in master thread, like constant expressions, null statements, simple declarations, etc., such construct can be xecuted in SPMD mode. llvm-svn: 346551
OpenPOWER on IntegriCloud