Commit Graph

43 Commits

Author SHA1 Message Date
Jon Chesterfield 72e8a4c45d [openmp][docs] Describe how the internal components are found
Add a FAQ entry about the names of openmp offloading components
and how they are searched for.

Reviewed By: jhuber6

Differential Revision: https://reviews.llvm.org/D109619
2021-09-30 22:05:12 +01:00
Joseph Huber f1c821fa85 [OpenMP] Add support for dynamic shared memory in new RTL
This patch adds support for using dynamic shared memory in the new
device runtime. The new function `__kmpc_get_dynamic_shared` will return a
pointer to the buffer of dynamic shared memory. Currently the amount of memory
allocated is set by an environment variable.

In the future this amount will be added to the amount used for the smart stack
which will be configured in a similar way.

Reviewed By: tianshilei1992

Differential Revision: https://reviews.llvm.org/D110006
2021-09-17 21:25:36 -04:00
Joseph Huber 7eb899cbcd [OpenMP] Add more verbose remarks for runtime folding
We peform runtime folding, but do not currently emit remarks when it is
performed. This is because it comes from the runtime library and is
beyond the users control. However, people may still wish to view  this
and similar information easily, so we can enable this behaviour using a
special flag to enable verbose remarks.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D109627
2021-09-10 17:36:06 -04:00
Jon Chesterfield f244af5c9f [openmp][amdgpu] Update SupportAndFAQ docs 2021-09-10 18:35:29 +01:00
Joel E. Denny 83ddfa0d22 [OpenMP][OpenACC] Implement `ompx_hold` map type modifier extension in Clang (1/2)
This patch implements Clang support for an original OpenMP extension
we have developed to support OpenACC: the `ompx_hold` map type
modifier.  The next patch in this series, D106510, implements OpenMP
runtime support.

Consider the following example:

```
 #pragma omp target data map(ompx_hold, tofrom: x) // holds onto mapping of x
 {
   foo(); // might have map(delete: x)
   #pragma omp target map(present, alloc: x) // x is guaranteed to be present
   printf("%d\n", x);
 }
```

The `ompx_hold` map type modifier above specifies that the `target
data` directive holds onto the mapping for `x` throughout the
associated region regardless of any `target exit data` directives
executed during the call to `foo`.  Thus, the presence assertion for
`x` at the enclosed `target` construct cannot fail.  (As usual, the
standard OpenMP reference count for `x` must also reach zero before
the data is unmapped.)

Justification for inclusion in Clang and LLVM's OpenMP runtime:

* The `ompx_hold` modifier supports OpenACC functionality (structured
  reference count) that cannot be achieved in standard OpenMP, as of
  5.1.
* The runtime implementation for `ompx_hold` (next patch) will thus be
  used by Flang's OpenACC support.
* The Clang implementation for `ompx_hold` (this patch) as well as the
  runtime implementation are required for the Clang OpenACC support
  being developed as part of the ECP Clacc project, which translates
  OpenACC to OpenMP at the directive AST level.  These patches are the
  first step in upstreaming OpenACC functionality from Clacc.
* The Clang implementation for `ompx_hold` is also used by the tests
  in the runtime implementation.  That syntactic support makes the
  tests more readable than low-level runtime calls can.  Moreover,
  upstream Flang and Clang do not yet support OpenACC syntax
  sufficiently for writing the tests.
* More generally, the Clang implementation enables a clean separation
  of concerns between OpenACC and OpenMP development in LLVM.  That
  is, LLVM's OpenMP developers can discuss, modify, and debug LLVM's
  extended OpenMP implementation and test suite without directly
  considering OpenACC's language and execution model, which can be
  handled by LLVM's OpenACC developers.
* OpenMP users might find the `ompx_hold` modifier useful, as in the
  above example.

See new documentation introduced by this patch in `openmp/docs` for
more detail on the functionality of this extension and its
relationship with OpenACC.  For example, it explains how the runtime
must support two reference counts, as specified by OpenACC.

Clang recognizes `ompx_hold` unless `-fno-openmp-extensions`, a new
command-line option introduced by this patch, is specified.

Reviewed By: ABataev, jdoerfert, protze.joachim, grokos

Differential Revision: https://reviews.llvm.org/D106509
2021-08-31 16:13:49 -04:00
Joseph Huber dead50d442 [OpenMP][NFC] Fix a few typos in OpenMP documentation
Summary:
Fixes some typos in the OpenMP documentation.
2021-07-26 16:03:47 -04:00
Joseph Huber 3817ba13ae [OpenMP] Add environment variables to change stack / heap size in the CUDA plugin
This patch adds support for two environment variables to configure the device.
``LIBOMPTARGET_STACK_SIZE`` sets the amount of memory in bytes that each thread
has for its stack. ``LIBOMPTARGET_HEAP_SIZE`` sets the amount of heap memory
that can be allocated using malloc / free on the device.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D106627
2021-07-22 21:40:02 -04:00
Joseph Huber b917a1d713 [OpenMP] Change AMDGCN to AMDGPU in the Cmake Module
Summary:
Change the name for targeting AMD offloading.
2021-07-20 12:52:53 -04:00
Joseph Huber 6242f9b966 [OpenMP][Documentation] Fix hyperlink location
Fixes the documentation hyperlinks not showing the header.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D106374
2021-07-20 12:42:32 -04:00
Tony Tye 038602139d [NFC] Correct documentation error in OpenMP release ReleaseNotes
Reviewed By: tianshilei1992

Differential Revision: https://reviews.llvm.org/D106330
2021-07-20 02:04:43 +00:00
Joseph Huber 1616407921 [OpenMP] Add remark documentation to the OpenMP webpage
This patch begins adding documentation for each remark emitted by
`openmp-opt`. This builds on the IDs introduced in D105939 so that users
can more easily identify each remark in the webpage.

Depends on D105939.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D106018
2021-07-16 14:09:43 -04:00
Joseph Huber 2190c48fde [OpenMP][Documentation] Add FAQ entry for CMake module
This patch adds documentation for using the CMake find module for OpenMP
target offloading provided by LLVM. It also removes the requirement for
AMD's architecture to be set as this isn't necessary for upstream LLVM.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D105051
2021-06-28 17:05:07 -04:00
Joseph Huber c9f3240c9d [OpenMP][Documentation] Add OpenMPOpt optimization section
Add some information about the optimizations currently provided by
OpenMPOpt. Every optimization performed should eventually be listed
here.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D105050
2021-06-28 17:05:03 -04:00
Joel E. Denny 48421ac441 [OpenMP] Improve ref count debug messages
For example, without this patch:

```
$ cat test.c
int main() {
  int x;
  #pragma omp target enter data map(alloc: x)
  #pragma omp target exit data map(release: x)
  ;
  return 0;
}
$ clang -fopenmp -fopenmp-targets=nvptx64-nvidia-cuda test.c
$ LIBOMPTARGET_DEBUG=1 ./a.out |& grep 'Creating\|Mapping exists'
Libomptarget --> Creating new map entry with HstPtrBegin=0x00007ffcace8e448, TgtPtrBegin=0x00007f12ef600000, Size=4, Name=unknown
Libomptarget --> Mapping exists with HstPtrBegin=0x00007ffcace8e448, TgtPtrBegin=0x00007f12ef600000, Size=4, updated RefCount=1
```

There are two problems in this example:

* `RefCount` is not reported when a mapping is created, but it might
  be 1 or infinite.  In this case, because it's created by `omp target
  enter data`, it's 1.  Seeing that would make later `RefCount`
  messages easier to understand.
* `RefCount` is still 1 at the `omp target exit data`, but it's
  reported as `updated`.  The reason it's still 1 is that, upon
  deletions, the reference count is generally not updated in
  `DeviceTy::getTgtPtrBegin`, where the report is produced.  Instead,
  it's zeroed later in `DeviceTy::deallocTgtPtr`, where it's actually
  removed from the mapping table.

This patch makes the following changes:

* Report the reference count when creating a mapping.
* Where an existing mapping is reported, always report a reference
  count action:
    * `update suppressed` when `UpdateRefCount=false`
    * `incremented`
    * `decremented`
    * `deferred final decrement`, which replaces the misleading
      `updated` in the above example
* Add comments to `DeviceTy::getTgtPtrBegin` to explain why it does
  not zero the reference count.  (Please advise if these comments miss
  the point.)
* For unified shared memory, don't report confusing messages like
  `RefCount=` or `RefCount= updated` given that reference counts are
  irrelevant in this case.  Instead, just report `for unified shared
  memory`.
* Use `INFO` not `DP` consistently for `Mapping exists` messages.
* Fix device table dumps to print `INF` instead of `-1` for an
  infinite reference count.

Reviewed By: jhuber6, grokos

Differential Revision: https://reviews.llvm.org/D104559
2021-06-23 09:57:19 -04:00
Asher Mancinelli 5c189d30e6 [OpenMP] Update FAQ for enabling cuda offloading
Add an FAQ entry and add a few lines to an existing one. Document
the use of `GCC_INSTALL_PREFIX` for pointing clang to correct
GCC installation for two-stage build.

Reviewed By: jhuber6

Differential Revision: https://reviews.llvm.org/D104474
2021-06-18 11:55:45 -06:00
Joseph Huber df965513a9 [OpenMP] Add an information flag for device data transfers
This patch adds an information flag that indicated when data is being copied to
and from the device. This will be helpful for finding redundant or unnecessary
data transfers in applications.

Reviewed By: jdoerfert, grokos

Differential Revision: https://reviews.llvm.org/D103927
2021-06-08 20:23:27 -04:00
Jon Chesterfield 25fe17d3c1 [libomptarget] Initial documentation on amdgpu offload
[libomptarget] Initial documentation on amdgpu offload

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D101927
2021-05-05 19:58:52 +01:00
Joseph Huber 077fe0f739 [OpenMP][Documentation] Add FAQ entry for dynamically linked libraries
Summary:
Add an FAW entry detailing the support for using dynamically linked libraries
with OpenMP Offloading
2021-04-26 14:21:17 -04:00
Joseph Huber 2b6f20082e [OpenMP] Add function for setting LIBOMPTARGET_INFO at runtime
Summary:
This patch adds a new runtime function __tgt_set_info_flag that allows the
user to set the information level at runtime without using the environment
variable. Using this will require an extern function, but will eventually be
added into an auxilliary library for OpenMP support functions.

This patch required moving the current InfoLevel to a global variable which must
be instantiated by each plugin.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D100774
2021-04-22 12:48:11 -04:00
Joseph Huber 83d4b2e2e0 [OpenMP] Add info for device table changes
Summary:
This patch adds a feature to print information whenever the host-device pointer
mapping table is changed by inserting or removing an entry. This introduces a
new bit field for LIBOMPTARGET_INFO at position 0x8.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D100600
2021-04-15 18:39:48 -04:00
Joseph Huber 29338459fb [OpenMP] Trim error messages in CUDA plugin
Summary:
Remove some of the error messages printed when the CUDA plugin fails. The current error messages can be confusing because they are the first error messages printed after the async stream finds an error. This means that the printed values aren't related to what caused the issue, but are simply the last asyncronous operation that succeeded on the device. Remove these as they can be misleading.

Reviewers: jdoerfert

Differential Revision: https://reviews.llvm.org/D99510
2021-03-29 12:20:19 -04:00
Joseph Huber ed8943c087 [OpenMP][NFC] Adding FAQ Entry for errors with static libraries 2021-02-02 10:50:22 -05:00
Shilei Tian 7bc31018f7 [OpenMP][NFC] Added release note for new `deviceRTLs` and hidden helper task
Added release note for new `deviceRTLs` and hidden helper task for LLVM
12.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D95584
2021-01-29 13:13:03 -05:00
Shilei Tian c571b16834 [OpenMP] Disabled profiling in `libomp` by default to unblock link errors
Link error occurred when time profiling in libomp is enabled by default
because `libomp` is assumed to be a C library but the dependence on
`libLLVMSupport` for profiling is a C++ library. Currently the issue blocks all
OpenMP tests in Phabricator.

This patch set a new CMake option `OPENMP_ENABLE_LIBOMP_PROFILING` to
enable/disable the feature. By default it is disabled. Note that once time
profiling is enabled for `libomp`, it becomes a C++ library.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D95585
2021-01-28 07:24:32 -05:00
Atmn Patel ec8f4a38c8 [OpenMP][Libomptarget] Introduce Remote Offloading Plugin
This introduces a remote offloading plugin for libomptarget. This
implementation relies on gRPC and protobuf, so this library will only
build if both libraries are available on the system. The corresponding
server is compiled to `openmp-offloading-server`.

This is a large change, but the only way to split this up is into RTL/server
but I fear that could introduce an inconsistency amongst them.

Ideally, tests for this should be added to the current ones that but that is
problematic for at least one reason. Given that libomptarget registers plugin
on a first-come-first-serve basis, if we wanted to offload onto a local x86
through a different process, then we'd have to either re-order the plugin list
in `rtl.cpp` (which is what I did locally for testing) or find a better
solution for runtime plugin registration in libomptarget.

Differential Revision: https://reviews.llvm.org/D95314
2021-01-26 15:33:38 -05:00
Kelvin Li 9d81073acb [OpenMP][Docs] Fix typos in FAQ (NFC) 2021-01-18 18:55:58 -05:00
Joseph Huber a957634942 [OpenMP] Add documentation for error messages and release notes
Add extra information to the runtime page describing the error messages and add information to the release notes for clang 12.0

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D94562
2021-01-13 11:00:41 -05:00
Johannes Doerfert 9ae171bcd3 [OpenMP][Docs] Add remarks intro section
Reviewed By: jhuber6

Differential Revision: https://reviews.llvm.org/D93735
2021-01-07 14:31:17 -06:00
Joseph Huber abb174bbc1 [OpenMP] Add example in Libomptarget Information docs
Add an example to the OpenMP Documentation on the LIBOMPTARGET_INFO environment variable

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D94246
2021-01-07 15:00:51 -05:00
Joseph Huber 631501b1f9 [OpenMP] Fixing typo on memory size in Documenation 2020-12-23 11:46:26 -05:00
Joseph Huber 6e60346495 [OpenMP] Fixing Typo in Documentation 2020-12-23 09:17:51 -05:00
Joseph Huber 1c19804ebf [OpenMP] Add OpenMP Documentation for Libomptarget environment variables
Add support to the OpenMP web pages for environment variables supported
by Libomptarget and their usage.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D93723
2020-12-22 17:41:27 -05:00
Johannes Doerfert 7b0f9dd79a [OpenMP][Docs] Fix Typo 2020-12-22 13:06:23 -06:00
Shilei Tian 1eb082c2ea [OpenMP][Docs] Fixed a typo in the doc that can mislead users to a CMake error
When setting `LLVM_ENABLE_RUNTIMES`, lower case word should be used;
otherwise, it can cause a CMake error that specific path is not found.

Reviewed By: ye-luo

Differential Revision: https://reviews.llvm.org/D93719
2020-12-22 14:05:58 -05:00
Johannes Doerfert 9cb748724e [OpenMP][Docs] Add FAQ entry about math and complex on GPUs
Reviewed By: tianshilei1992

Differential Revision: https://reviews.llvm.org/D93718
2020-12-22 13:05:04 -06:00
Shilei Tian 612ddc3117 [OpenMP][Docs] Updated the faq about building an OpenMP offloading capable compiler
After some issues about building runtimes along with LLVM were fixed,
building an OpenMP offloading capable compiler is pretty simple. This patch updates
the FAQ part in the doc.

Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D93671
2020-12-22 13:14:53 -05:00
Johannes Doerfert 994bb6eb7d [OpenMP][NFC] Provide a new remark and documentation
If a GPU function is externally reachable we give up trying to find the
(unique) kernel it is called from. This can hinder optimizations. Emit a
remark and explain mitigation strategies.

Reviewed By: tianshilei1992

Differential Revision: https://reviews.llvm.org/D93439
2020-12-17 14:38:26 -06:00
Giorgis Georgakoudis 18dff28958 [OpenMP] Add doxygen generation for the runtime
Reviewed By: jdoerfert

Differential Revision: https://reviews.llvm.org/D92779
2020-12-08 16:20:45 -08:00
Johannes Doerfert 227c8ff189 [OpenMP][Docs] Add more content, call coordinates, FAQ entries, links 2020-11-25 11:52:35 -06:00
Johannes Doerfert 30e818db91 [OpenMP][Docs] Structure and content for the OpenMP documentation
This adds some initial content as well as structure to the new OpenMP
Sphinx documentation hosted at http://openmp.llvm.org/docs/ .

The content contains some useful links but most pages are still empty.

This uses a "custom" theme which is a copy of the default "agogo" one
with minor modifications to get a nicer table of content in the sidebar.
This way we can also adjust the theme as we go.

Reviewed By: jhuber6, JonChesterfield

Differential Revision: https://reviews.llvm.org/D90256
2020-10-30 01:31:48 -05:00
Joseph Huber 0103df7903 [OpenMP] Add Missing _static Director for OpenMP Documentation
Summary:
Adding a missing directory needed for generating Sphinx documentation without
errors. Directory current contains a placeholder image just to populate the
directory.
2020-09-27 15:35:47 -04:00
Joseph Huber 1c4c21489f [OpenMP] Initial Support for OpenMP Webpage Documentation
Summary:
Adding support for generated html documentation for OpenMP. Changing
Cmake files to build the documentation and adding the base templates for
future documentation to be added.

Reviewers: jdoerfert

Subscribers: aaron.ballman arphaman guansong mgorny openmp-commits sstefan1 yaxunl

Tags: #OpenMP

Differential Revision: https://reviews.llvm.org/D87797
2020-09-18 16:32:22 -04:00
Joachim Protze 73b7ff4e16 [OpenMP] NFC: Create OpenMP release notes file 2020-06-24 13:42:32 +02:00