Commit Graph

376 Commits

Author SHA1 Message Date
Richard Smith c70f1d63f8 [c++20] P0515R3: Parsing support and basic AST construction for operator <=>.
Adding the new enumerator forced a bunch more changes into this patch than I
would have liked. The -Wtautological-compare warning was extended to properly
check the new comparison operator, clang-format needed updating because it uses
precedence levels as weights for determining where to break lines (and several
operators increased their precedence levels with this change), thread-safety
analysis needed changes to build its own IL properly for the new operator.

All "real" semantic checking for this operator has been deferred to a future
patch. For now, we use the relational comparison rules and arbitrarily give
the builtin form of the operator a return type of 'void'.

llvm-svn: 320707
2017-12-14 15:16:18 +00:00
Alexey Bataev fbe17fb8a5 [OPENMP] Initial codegen for `target teams distribute simd` directive.
Host + generic device codegen for `target teams distribute simd`
directive.

llvm-svn: 320608
2017-12-13 19:45:06 +00:00
Alexey Bataev e83b3e89e6 [OPENMP] Simplify codegen for loop iteration variables in loop preamble.
Initial patch could cause trouble in the optimized code because of the
incorrectly generated lifetime intrinsics.

llvm-svn: 320191
2017-12-08 20:18:58 +00:00
Alexey Bataev dfa430f694 [OPENMP] Initial codegen for `target teams distribute` directive.
Host + default devices codegen for `target teams distribute` directive.

llvm-svn: 320149
2017-12-08 15:03:50 +00:00
Alexey Bataev 999277ad22 [OPENMP] Initial codegen for `teams distribute simd` directive.
Host + default devices codegen for `teams distribute simd` directive.

llvm-svn: 319896
2017-12-06 14:31:09 +00:00
Alexey Bataev c2e88a8a6b [OPENMP] Fix PR35486: crash when collapsing loops with dependent iteration 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
2017-12-04 21:30:42 +00:00
Carlo Bertolli 56a2aa4ddc [OpenMP] Initial implementation of code generation for pragma 'teams distribute parallel for simd' on host
https://reviews.llvm.org/D40795

This includes regression tests for all associated clauses.

llvm-svn: 319696
2017-12-04 20:57:19 +00:00
Alexey Bataev 617db5f822 [OPENMP] Codegen for `distribute simd` directive.
Initial codegen support for `distribute simd` directive.

llvm-svn: 319661
2017-12-04 15:38:33 +00:00
Alexey Bataev 95c6dd49a0 [OPENMP] General improvement of handling of `teams distribute`
directive, NFC.

Some general improvements in support of `teams distribute` directive.

llvm-svn: 319320
2017-11-29 15:14:16 +00:00
Alexey Bataev 0b49f9e489 [OPENMP] Codegen for `distribute parallel for simd` directive.
Initial codegen for `#pragma omp distribute parallel for simd` directive
and its clauses.

llvm-svn: 319079
2017-11-27 19:38:58 +00:00
Alexey Bataev 10a5431239 [OPENMP] Improve handling of cancel directives in target-based
constructs, NFC.

Improved handling of cancel|cancellation point directives inside
target-based for directives.

llvm-svn: 319046
2017-11-27 16:54:08 +00:00
Alexey Bataev 16e798873e [OPENMP] Add support for cancel constructs in `target teams distribute
parallel for`.

Add support for cancel/cancellation point directives inside `target
teams distribute parallel for` directives.

llvm-svn: 318881
2017-11-22 21:12:03 +00:00
Alexey Bataev dcb4b8fbc1 [OPENMP] Add support for cancel constructs in [teams] distribute
parallel for directives.

Added codegen/sema support for cancel constructs in [teams] distribute
parallel for directives.

llvm-svn: 318872
2017-11-22 20:19:50 +00:00
Alexey Bataev b45d43c397 [OPENMP] Do not mark captured variables as artificial in debug info.
Captured variables should not be marked as artificial parameters in
outlined functions in debug info.

llvm-svn: 318843
2017-11-22 16:02:03 +00:00
Alexey Bataev f9fc42e50b [OPENMP] Codegen for `target teams` directive.
Added codegen of the clauses for `target teams` directive.

llvm-svn: 318834
2017-11-22 14:25:55 +00:00
Alexey Bataev 7828b25251 [OPENMP] Initial support for asynchronous data update, NFC.
OpenMP 5.0 introduces asynchronous data update/dependecies clauses on
target data directives. Patch adds initial support for outer task
regions to use task-based codegen for future async target data
directives.

llvm-svn: 318781
2017-11-21 17:08:48 +00:00
Carlo Bertolli 62fae15600 [OpenMP] Initial implementation of code generation for pragma 'teams 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
2017-11-20 20:46:39 +00:00
Hans Wennborg 989a65cd29 Fix some -Wunused-variable warnings
llvm-svn: 318578
2017-11-18 00:49:18 +00:00
Alexey Bataev f836537516 [OPENMP] Codegen for `target simd` construct.
Added codegen support for `target simd` directive.

llvm-svn: 318536
2017-11-17 17:57:25 +00:00
Alexey Bataev 2139ed638b [OPENMP] Add support for cancelling inside target parallel for
directive.

Added missed support for cancelling of target parallel for construct.

llvm-svn: 318434
2017-11-16 18:20:21 +00:00
Alexey Bataev 5d7edca316 [OPENMP] Codegen for `#pragma omp target parallel for simd`.
Added codegen for `#pragma omp target parallel for simd` and clauses.

llvm-svn: 317813
2017-11-09 17:32:15 +00:00
Alexey Bataev fb0ebecf0e [OPENMP] Codegen for `#pragma omp target parallel for`.
llvm-svn: 317719
2017-11-08 20:16:14 +00:00
Ivan A. Kosarev 9f9d157517 [CodeGen] Generate TBAA info for reference loads
Differential Revision: https://reviews.llvm.org/D39177

llvm-svn: 316896
2017-10-30 11:49:31 +00:00
Alexey Bataev 1b48c5e56b [OPENMP] Fix PR35013: Fix passing VLAs captures to outlined functions.
Fixed passing of VLAs and variably-modified types to outlined functions.
Synchronized passing with the types codegen.

llvm-svn: 316488
2017-10-24 19:52:31 +00:00
Jonas Hahnfeld 4525c82428 [OpenMP] Avoid VLAs for some reductions on array sections
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
2017-10-23 19:01:35 +00:00
Jonas Hahnfeld c95a6985bd Revert "[OpenMP] Avoid VLAs for some reductions on array sections"
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
2017-10-20 20:16:17 +00:00
Jonas Hahnfeld b6229be460 [OpenMP] Avoid VLAs for some reductions on array sections
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
2017-10-20 19:40:40 +00:00
Alexey Bataev 7ba57afd6e [OPENMP] Fix capturing of boolean variables in debug mode.
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
2017-10-17 16:47:34 +00:00
Alexey Bataev 2a0c4f57dd [OPENMP] Add default codegen|tests for 'target parallel for[ simd]'
constructs.

Added default codegen for 'target parallel for' construct + tests for
default codegen of 'target parallel for[ simd]' constructs.

llvm-svn: 315315
2017-10-10 14:14:43 +00:00
Ivan A. Kosarev 5f8c0ca53d [CodeGen] Do not construct complete LValue base info in trivial cases
Besides obvious code simplification, avoiding explicit creation
of LValueBaseInfo objects makes it easier to make TBAA
information to be part of such objects.

This is part of D38126 reworked to be a separate patch to
simplify review.

Differential Revision: https://reviews.llvm.org/D38695

llvm-svn: 315289
2017-10-10 09:39:32 +00:00
Carlo Bertolli ba1487ba69 [OpenMP] Initial implementation of teams distribute code generation
https://reviews.llvm.org/D38371

This patch implements codegen for the combined 'teams distribute" OpenMP pragma and adds regression tests for all its clauses.

llvm-svn: 314905
2017-10-04 14:12:09 +00:00
Alexey Bataev 931e19bf51 [OPENMP] Capture argument of `device` clause for target-based
directives.

The argument of the `device` clause in target-based executable
directives must be captured to support codegen for the `target`
directives with the `depend` clauses.

llvm-svn: 314686
2017-10-02 16:32:39 +00:00
Alexey Bataev f47c4b4184 [OPENMP] Generate implicit map|firstprivate clauses for target-based
directives.

If the variable is used in the target-based region but is not found in
any private|mapping clause, then generate implicit firstprivate|map
clauses for these implicitly mapped variables.

llvm-svn: 314205
2017-09-26 13:47:31 +00:00
Alexey Bataev f43f714213 [OPENMP] Fix for PR33922: New ident_t flags for
__kmpc_for_static_fini().

Added special flags for calls of __kmpc_for_static_fini(), like previous
ly for __kmpc_for_static_init(). Added flag OMP_IDENT_WORK_DISTRIBUTE
for distribute cnstruct, OMP_IDENT_WORK_SECTIONS for sections-based
  constructs and OMP_IDENT_WORK_LOOP for loop-based constructs in
  location flags.

llvm-svn: 312642
2017-09-06 16:17:35 +00:00
Alexey Bataev 6a71f364f1 [OPENMP] Fix for PR34014: OpenMP 4.5: Target construct in static method
of class fails to map class static variable.

If the global variable is captured and it has several redeclarations,
sometimes it may lead to a compiler crash. Patch fixes this by working
only with canonical declarations.

llvm-svn: 311479
2017-08-22 17:54:52 +00:00
Alexey Bataev 8c3edfef6b [OPENMP] Fix for PR28581: OpenMP linear clause - wrong results.
If worksharing construct has at least one linear item, an implicit
synchronization point must be emitted to avoid possible conflict with
the loading/storing values to the original variables. Added implicit
barrier if the linear item is found before actual start of the
worksharing construct.

llvm-svn: 311013
2017-08-16 15:58:46 +00:00
Alexey Bataev 0f87dbee4e [OPENMP] Fix for PR33922: New ident_t flags for
__kmpc_for_static_init().

OpenMP 5.0 will include OpenMP Tools interface that requires distinguishing different worksharing constructs.

Since the same entry point (__kmp_for_static_init(ident_t *loc,
kmp_int32 global_tid,........)) is called in case static
loop/sections/distribute it is suggested using 'flags' field of the
ident_t structure to pass the type of the construct.

llvm-svn: 310865
2017-08-14 17:56:13 +00:00
Alexey Bataev 6e01dc1b84 [OPENMP][DEBUG] Fix for PR33676: Debug info for OpenMP region is broken.
After some changes in clang/LLVM debug info for task-based regions was
not generated at all. Patch fixes this problem.

llvm-svn: 310850
2017-08-14 16:03:47 +00:00
Alexey Bataev 3c595a6b2c [OPENMP] Generalization of calls of the outlined functions.
General improvement of the outlined functions calls.

llvm-svn: 310840
2017-08-14 15:01:03 +00:00
Alexey Bataev e754b18f5e [OPENMP] Emit non-debug version of outlined functions with original
name.

If the host code is compiled with the debug info, while the target
without, there is a problem that the compiler is unable to find the
debug wrapper. Patch fixes this problem by emitting special name for the
debug version of the code.

llvm-svn: 310511
2017-08-09 19:38:53 +00:00
Alexey Bataev 3b8d5586ec [OPENMP][DEBUG] Set proper address space info if required by target.
Arguments, passed to the outlined function, must have correct address
space info for proper Debug info support. Patch sets global address
space for arguments that are mapped and passed by reference.

Also, cuda-gdb does not handle reference types correctly, so reference
arguments are represented as pointers.

llvm-svn: 310387
2017-08-08 18:04:06 +00:00
Alexey Bataev 4aa19052f3 Revert "[OPENMP][DEBUG] Set proper address space info if required by target."
This reverts commit r310377.

llvm-svn: 310379
2017-08-08 16:45:36 +00:00
Alexey Bataev 5a497136be [OPENMP][DEBUG] Set proper address space info if required by target.
Arguments, passed to the outlined function, must have correct address
space info for proper Debug info support. Patch sets global address
space for arguments that are mapped and passed by reference.

Also, cuda-gdb does not handle reference types correctly, so reference
arguments are represented as pointers.

llvm-svn: 310377
2017-08-08 16:29:11 +00:00
Alexey Bataev 6a824b9a45 Revert "[OPENMP][DEBUG] Set proper address space info if required by target."
This reverts commit r310360.

llvm-svn: 310364
2017-08-08 14:44:43 +00:00
Alexey Bataev 59b81e51d3 [OPENMP][DEBUG] Set proper address space info if required by target.
Arguments, passed to the outlined function, must have correct address
space info for proper Debug info support. Patch sets global address
space for arguments that are mapped and passed by reference.

Also, cuda-gdb does not handle reference types correctly, so reference
arguments are represented as pointers.

llvm-svn: 310360
2017-08-08 14:25:14 +00:00
Alexey Bataev d90ec748a8 Revert "[OPENMP][DEBUG] Set proper address space info if required by target."
This reverts commit r310104.

llvm-svn: 310135
2017-08-04 21:27:11 +00:00
Alexey Bataev efd884d393 Revert "[OPENMP] Fix for pacify buildbots, NFC."
This reverts commit r310120.

llvm-svn: 310134
2017-08-04 21:26:25 +00:00
Alexey Bataev e09a774bea [OPENMP] Fix for pacify buildbots, NFC.
llvm-svn: 310120
2017-08-04 20:29:52 +00:00
Alexey Bataev be83fad57e [OPENMP][DEBUG] Set proper address space info if required by target.
Arguments, passed to the outlined function, must have correct address
space info for proper Debug info support. Patch sets global address
space for arguments that are mapped and passed by reference.

Also, cuda-gdb does not handle reference types correctly, so reference
arguments are represented as pointers.

llvm-svn: 310104
2017-08-04 19:46:10 +00:00
Alexey Bataev 2c7eee5b84 [OPENMP] Unify generation of outlined function calls.
llvm-svn: 310098
2017-08-04 19:10:54 +00:00