summaryrefslogtreecommitdiffstats
path: root/clang/test/OpenMP
Commit message (Collapse)AuthorAgeFilesLines
...
* [OPENMP]Improve detection of number of teams, threads in targetAlexey Bataev2019-04-1013-47/+47
| | | | | | | | | | regions. Added more complex analysis for number of teams and number of threads in the target regions, also merged related common code between CGOpenMPRuntime and CGOpenMPRuntimeNVPTX classes. llvm-svn: 358126
* Don't emit an unreachable return block.John McCall2019-04-101-1/+1
| | | | | | Patch by Brad Moody. llvm-svn: 358104
* [OPENMP]Allow allocate directive on parameters.Alexey Bataev2019-04-091-0/+14
| | | | | | Patch allows to use allocate directives on the function parameters. llvm-svn: 358016
* [OPENMP] Sync __kmpc_alloc/_kmpc_free function with the runtime.Alexey Bataev2019-04-086-87/+102
| | | | | | | Functions __kmpc_alloc/__kmpc_free are updated with the new interfaces. Patch synchronizes the compiler with the runtime. llvm-svn: 357933
* [OPENMP][NVPTX]Fixed processing of memory management directives.Alexey Bataev2019-04-081-1/+32
| | | | | | | | | | | Added special processing of the memory management directives/clauses for NVPTX target. For private locals, omp_default_mem_alloc and omp_thread_mem_alloc result in allocation in local memory. omp_const_mem_alloc allocates const memory, omp_teams_mem_alloc allocates shared memory, and omp_cgroup_mem_alloc and omp_large_cap_mem_alloc allocate global memory. llvm-svn: 357923
* [OPENMP]Add codegen for task reduction vars with allocate clause, NFC.Alexey Bataev2019-04-042-2/+22
| | | | | | Added test for the task reduction variables with the allocate clause. llvm-svn: 357717
* [OPENMP]Add codegen for linear vars with allocate clause, NFC.Alexey Bataev2019-04-041-3/+16
| | | | | | Added test for the linear variables with the allocate clause. llvm-svn: 357712
* [OPENMP]Fix lookup of the user-defined reductions in C.Alexey Bataev2019-04-041-0/+13
| | | | | | Fixed the regression of the lookup of user-defined reductions for C. llvm-svn: 357708
* [OPENMP]Add codegen for reduction vars with allocate clause, NFC.Alexey Bataev2019-04-031-3/+15
| | | | | | Added test for the reduction variables with the allocate clause. llvm-svn: 357629
* [OPENMP]Add codegen for lastprivate vars with allocate clause, NFC.Alexey Bataev2019-04-031-11/+25
| | | | | | Added test for the lastprivatized variables with the allocate clause. llvm-svn: 357625
* [OPENMP]Add codegen for firstprivate vars with allocate clause.Alexey Bataev2019-04-031-4/+58
| | | | | | | Added codegen/test for the firstprivatized variables with the allocate clause. llvm-svn: 357617
* [OPENMP]Add codegen for private vars with allocate clause.Alexey Bataev2019-04-021-2/+18
| | | | | | | Added codegen/test for the privatized variables with the allocate clause. llvm-svn: 357514
* [OPENMP]Fix mapping of the pointers captured by reference.Alexey Bataev2019-04-021-13/+25
| | | | | | | | If the pointer is captured by reference, it must be mapped as _PTR_AND_OBJ kind of mapping to correctly translate the pointer address on the device. llvm-svn: 357488
* [CodeGen] Generate follow-up metadata for loops with more than one ↵Michael Kruse2019-04-011-3/+3
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | transformation. Before this patch, CGLoop would dump all transformations for a loop into a single LoopID without encoding any order in which to apply them. rL348944 added the possibility to encode a transformation order using followup-attributes. When a loop has more than one transformation, use the follow-up attribute define the order in which they are applied. The emitted order is the defacto order as defined by the current LLVM pass pipeline, which is: LoopFullUnrollPass LoopDistributePass LoopVectorizePass LoopUnrollAndJamPass LoopUnrollPass MachinePipeliner This patch should therefore not change the assembly output, assuming that all explicit transformations can be applied, and no implicit transformations in-between. In the former case, WarnMissedTransformationsPass should emit a warning (except for MachinePipeliner which is not implemented yet). The latter could be avoided by adding 'llvm.loop.disable_nonforced' attributes. Because LoopUnrollAndJamPass processes a loop nest, generation of the MDNode is delayed to after the inner loop metadata have been processed. A temporary LoopID is therefore used to annotate instructions and RAUW'ed by the actual LoopID later. Differential Revision: https://reviews.llvm.org/D57978 llvm-svn: 357415
* [OPENMP]Allocate clause allocator in target region.Alexey Bataev2019-04-011-0/+6
| | | | | | | | | | According to OpenMP 5.0, 2.11.4 allocate Clause, Restrictions, allocate clauses that appear on a target construct or on constructs in a target region must specify an allocator expression unless a requires directive with the dynamic_allocators clause is present in the same compilation unit. Patch adds a check for this restriction. llvm-svn: 357412
* [OPENMP] Check that allocated variables are used in private clauses.Alexey Bataev2019-04-011-0/+2
| | | | | | | | | | According to OpenMP 5.0 standard, 2.11.4 allocate Clause, Restrictions, For any list item that is specified in the allocate clause on a directive, a data-sharing attribute clause that may create a private copy of that list item must be specified on the same directive. Patch adds the checks for this restriction. llvm-svn: 357390
* [OPENMP]Add check for undefined behavior with thread allocators onAlexey Bataev2019-03-2856-115/+619
| | | | | | | | | | | | target and task-based directives. According to OpenMP 5.0, 2.11.4 allocate Clause, Restrictions, For task, taskloop or target directives, allocation requests to memory allocators with the trait access set to thread result in unspecified behavior. Patch introduces a check for omp_thread_mem_alloc predefined allocator on target- and trask-based directives. llvm-svn: 357205
* [OPENMP]Initial support for 'allocate' clause.Alexey Bataev2019-03-27181-269/+413
| | | | | | Added parsing/sema analysis of the allocate clause. llvm-svn: 357068
* [OPENMP]Allow no allocator clause in target regions with requiresAlexey Bataev2019-03-221-7/+12
| | | | | | | | | | | | dynamic_allocators. According to the OpenMP 5.0, 2.11.3 allocate Directive, Restrictions, allocate directives that appear in a target region must specify an allocator clause unless a requires directive with the dynamic_allocators clause is present in the same compilation unit. Patch adds a check for a presence of the requires directive with the dynamic_allocators clause. llvm-svn: 356758
* [OPENMP]Emit error message for allocate directive without allocatorAlexey Bataev2019-03-222-1/+81
| | | | | | | | | | | clause in target region. According to the OpenMP 5.0, 2.11.3 allocate Directive, Restrictions, allocate directives that appear in a target region must specify an allocator clause unless a requires directive with the dynamic_allocators clause is present in the same compilation unit. llvm-svn: 356752
* [OPENMP] Simplify codegen for allocate directive on local variables.Alexey Bataev2019-03-212-12/+7
| | | | | | | Simplified codegen for the allocate directive for local variables, initial implementation of the codegen for NVPTX target. llvm-svn: 356710
* [OPENMP]Codegen support for allocate directive on global variables.Alexey Bataev2019-03-211-0/+71
| | | | | | | | | For the global variables the allocate directive must specify only the predefined allocator. This allocator must be translated into the correct form of the address space for the targets that support different address spaces. llvm-svn: 356702
* [OPENMP]Improve detection of omp_allocator_handle_t type and predefinedAlexey Bataev2019-03-201-2/+8
| | | | | | | | | | | allocators. It is better to deduce omp_allocator_handle_t type from the predefined allocators, because omp.h header might not define it explicitly. Plus, it allows to identify the predefined allocators correctly when trying to build the allcoator for the global variables. llvm-svn: 356607
* [OPENMP]Warn if the different allocator is used for the variable.Alexey Bataev2019-03-193-11/+23
| | | | | | | | If the allocator was specified for the variable and next one is found with the different allocator, the warning is emitted, and the allocator is ignored. llvm-svn: 356513
* [OPENMP]Check that global vars require predefined allocator.Alexey Bataev2019-03-191-1/+3
| | | | | | | | | | According to OpenMP, 2.11.3 allocate Directive, Restrictions, C / C++, if a list item has a static storage type, the allocator expression in the allocator clause must be a constant expression that evaluates to one of the predefined memory allocator values. Added check for this restriction. llvm-svn: 356496
* [OPENMP] Codegen for local variables with the allocate pragma.Alexey Bataev2019-03-192-8/+114
| | | | | | | | Added initial codegen for the local variables with the #pragma omp allocate directive. Instead of allocating the variables on the stack, __kmpc_alloc|__kmpc_free functions are used for memory (de-)allocation. llvm-svn: 356472
* [OPENMP] Set scheduling for doacross loops as schedule, 1.Alexey Bataev2019-03-182-3/+3
| | | | | | | The default scheduling for doacross loops is changed from static to static, 1. llvm-svn: 356388
* [OPENMP]Fix crash for the ordered(n) clause.Alexey Bataev2019-03-141-1/+12
| | | | | | | | If the doacross lop construct is used and the loop counter is declare outside of the loop, the compiler might crash trying to get the address of the loop counter. Patch fixes this problem. llvm-svn: 356198
* [OPENMP]Fix PR37283: Assertion failure on openmp task with by referenceAlexey Bataev2019-03-131-2/+3
| | | | | | | | | | array. If the firstprivate variable is a reference, we may incorrectly classify the kind of the private copy. Use the type of the private copy instead of the original shared variable. llvm-svn: 356098
* [OPENMP]Disable ADL in C for user-defined reductions.Alexey Bataev2019-03-131-0/+10
| | | | | | C does not support ADL, disable it for C to prevent compiler crash. llvm-svn: 356089
* [OPENMP][NVPTX]Fix PR40893: Size doesn't match forAlexey Bataev2019-03-131-1/+1
| | | | | | | | | | | '_openmp_teams_reductions_buffer_$_. nvlink does not handle weak linkage correctly, same symbols with the different sizes are reported as erroneous though the largest size must be chosen instead. Patch fixes this problem by using Internal linkage instead of the Common. llvm-svn: 356072
* [OPENMP]Allow to redefine entry for the variables definitions.Alexey Bataev2019-03-121-2/+14
| | | | | | | | | | | If the variable was declared and marked as declare target, a new offload entry with size 0 is created. But if later a definition is created and marked as declare target, this definition is not added to the entry set and the definition remains not mapped to the target. Patch fixes this problem allowing to redefine the size and linkage for previously registered declaration. llvm-svn: 355960
* [OPENMP 5.0]Initial support for 'allocator' clause.Alexey Bataev2019-03-122-0/+116
| | | | | | | Added parsing/sema analysis/serialization/deserialization for the 'allocator' clause of the 'allocate' directive. llvm-svn: 355952
* [OPENMP]Fix codegen for declare target link in target regions.Alexey Bataev2019-03-111-3/+7
| | | | | | | | If the declare target link global is used in the target region indirectly (used in the inner parallel, teams, etc. regions), we may miss this variable and it leads to incorrect codegen. llvm-svn: 355858
* [OPENMP]Remove debug service variable.Alexey Bataev2019-03-081-2/+1
| | | | | | Removed not required service variable for the debug info. llvm-svn: 355729
* [OPENMP 5.0]Add initial support for 'allocate' directive.Alexey Bataev2019-03-072-0/+228
| | | | | | | Added parsing/sema analysis/serialization/deserialization support for 'allocate' directive. llvm-svn: 355614
* [OPENMP]Target region: emit const firstprivates as globals with constantAlexey Bataev2019-03-052-59/+73
| | | | | | | | | | | | memory. If the variable with the constant non-scalar type is firstprivatized in the target region, the local copy is created with the data copying. Instead, we allocate the copy in the constant memory and avoid extra copying in the outlined target regions. This global copy is used in the target regions without loss of the performance. llvm-svn: 355418
* [OPENMP]Delay emission of the error for unsupported types.Alexey Bataev2019-02-272-0/+111
| | | | | | | If the type is unsupported on the device side, it still must be emitted, but we should emit errors for operations with such types. llvm-svn: 355027
* [OPENMP][CUDA]Do not emit warnings for variables in late-reported asmAlexey Bataev2019-02-261-3/+3
| | | | | | | | | | | | statements. If the assembler instruction is not generated and the delayed diagnostic is emitted, we may end up with extra warning message for variables used in the asm statement. Since the asm statement is not built, the variables may be left non-referenced and it may produce a warning about a use of the non-initialized variables. llvm-svn: 354928
* [OPENMP]Delay emission for unsupported va_arg expression.Alexey Bataev2019-02-261-0/+40
| | | | | | | If the OpenMP device is NVPTX and va_arg is used, delay emission of the error for va_arg unless it is used in the device code. llvm-svn: 354925
* [OpenMP 5.0] Parsing/sema support for from clause with mapper modifier.Michael Kruse2019-02-255-3/+49
| | | | | | | | | | | | | | | | | | | | This patch implements the parsing and sema support for the OpenMP 'from'-clause with potential user-defined mappers attached. User-defined mappers are a new feature in OpenMP 5.0. A 'from'-clause can have an explicit or implicit associated mapper, which instructs the compiler to generate and use customized mapping functions. An example is shown below: struct S { int len; int *d; }; #pragma omp declare mapper(id: struct S s) map(s, s.d[0:s.len]) struct S ss; #pragma omp target update from(mapper(id): ss) // use the mapper with name 'id' to map ss from device Contributed-by: Lingda Li <lildmh@gmail.com> Differential Revision: https://reviews.llvm.org/D58638 llvm-svn: 354817
* [OpenMP 5.0] Parsing/sema support for to clause with mapper modifier.Michael Kruse2019-02-225-8/+56
| | | | | | | | | | | | | | | | | | | This patch implements the parsing and sema support for OpenMP to clause with potential user-defined mappers attached. User defined mapper is a new feature in OpenMP 5.0. A to/from clause can have an explicit or implicit associated mapper, which instructs the compiler to generate and use customized mapping functions. An example is shown below: struct S { int len; int *d; }; #pragma omp declare mapper(id: struct S s) map(s, s.d[0:s.len]) struct S ss; #pragma omp target update to(mapper(id): ss) // use the mapper with name 'id' to map ss to device Contributed-by: <lildmh@gmail.com> Differential Revision: https://reviews.llvm.org/D58523 llvm-svn: 354698
* [OPENMP] Delayed diagnostics for VLA support.Alexey Bataev2019-02-221-1/+4
| | | | | | Generalized processing of the deferred diagnostics for OpenMP/CUDA code. llvm-svn: 354690
* Revert "[OPENMP] Delayed diagnostics for VLA support."Alexey Bataev2019-02-221-4/+1
| | | | | | | This reverts commit r354679 to fix the problem with the Windows buildbots llvm-svn: 354680
* [OPENMP] Delayed diagnostics for VLA support.Alexey Bataev2019-02-221-1/+4
| | | | | | Generalized processing of the deferred diagnostics for OpenMP/CUDA code. llvm-svn: 354679
* [OPENMP] Delay emission of the asm target-specific error messages.Alexey Bataev2019-02-201-0/+118
| | | | | | | | | | | | | | | | | Summary: Added the ability to emit target-specific builtin assembler error messages only in case if the function is really is going to be emitted for the device. Reviewers: rjmccall Subscribers: guansong, jdoerfert, cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D58243 llvm-svn: 354486
* [OPENMP][NVPTX]Use faster teams reduction algorithm.Alexey Bataev2019-02-201-26/+940
| | | | | | | A faster way to reduce the values in teams reductions was found, the codegen is updated to use this faster algorithm and new runtime functions. llvm-svn: 354479
* [OpenMP 5.0] Parsing/sema support for map clause with mapper modifier.Michael Kruse2019-02-1915-29/+176
| | | | | | | | | | | | | | | | | | | This patch implements the parsing and sema support for OpenMP map clauses with potential user-defined mapper attached. User defined mapper is a new feature in OpenMP 5.0. A map clause can have an explicit or implicit associated mapper, which instructs the compiler to generate extra data mapping. An example is shown below: struct S { int len; int *d; }; #pragma omp declare mapper(id: struct S s) map(s, s.d[0:s.len]) struct S ss; #pragma omp target map(mapper(id) tofrom: ss) // use the mapper with name 'id' to map ss Contributed-by: Lingda Li <lildmh@gmail.com> Differential Revision: https://reviews.llvm.org/D58074 llvm-svn: 354347
* PR40642: Fix determination of whether the final statement of a statementRichard Smith2019-02-156-20/+26
| | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | | expression is a discarded-value expression. Summary: We used to get this wrong in three ways: 1) During parsing, an expression-statement followed by the }) ending a statement expression was always treated as producing the value of the statement expression. That's wrong for ({ if (1) expr; }) 2) During template instantiation, various kinds of statement (most statements not appearing directly in a compound-statement) were not treated as discarded-value expressions, resulting in missing volatile loads (etc). 3) In all contexts, an expression-statement with attributes was not treated as producing the value of the statement expression, eg ({ [[attr]] expr; }). Also fix incorrect enforcement of OpenMP rule that directives can "only be placed in the program at a position where ignoring or deleting the directive would result in a program with correct syntax". In particular, a label (be it goto, case, or default) should not affect whether directives are permitted. Reviewers: aaron.ballman, rjmccall Subscribers: cfe-commits Tags: #clang Differential Revision: https://reviews.llvm.org/D57984 llvm-svn: 354090
* [OPENMP]Delay emission of the error messages for the exceptions.Alexey Bataev2019-02-081-0/+13
| | | | | | | | | | Fixed diagnostic emission for the exceptions support in case of the compilation of OpenMP code for the devices. From now on, it uses delayed diagnostics mechanism, previously used for CUDA only. It allow to diagnose not allowed used of exceptions only in functions that are going to be codegen'ed. llvm-svn: 353542
OpenPOWER on IntegriCloud