1 \input texinfo @c -*-texinfo-*-
4 @setfilename libgomp.info
10 Copyright @copyright{} 2006-2024 Free Software Foundation, Inc.
12 Permission is granted to copy, distribute and/or modify this document
13 under the terms of the GNU Free Documentation License, Version 1.3 or
14 any later version published by the Free Software Foundation; with the
15 Invariant Sections being ``Funding Free Software'', the Front-Cover
16 texts being (a) (see below), and with the Back-Cover Texts being (b)
17 (see below). A copy of the license is included in the section entitled
18 ``GNU Free Documentation License''.
20 (a) The FSF's Front-Cover Text is:
24 (b) The FSF's Back-Cover Text is:
26 You have freedom to copy and modify this GNU Manual, like GNU
27 software. Copies published by the Free Software Foundation raise
28 funds for GNU development.
32 @dircategory GNU Libraries
34 * libgomp: (libgomp). GNU Offloading and Multi Processing Runtime Library.
37 This manual documents libgomp, the GNU Offloading and Multi Processing
38 Runtime library. This is the GNU implementation of the OpenMP and
39 OpenACC APIs for parallel and accelerator programming in C/C++ and
42 Published by the Free Software Foundation
43 51 Franklin Street, Fifth Floor
44 Boston, MA 02110-1301 USA
50 @setchapternewpage odd
53 @title GNU Offloading and Multi Processing Runtime Library
54 @subtitle The GNU OpenMP and OpenACC Implementation
56 @vskip 0pt plus 1filll
57 @comment For the @value{version-GCC} Version*
59 Published by the Free Software Foundation @*
60 51 Franklin Street, Fifth Floor@*
61 Boston, MA 02110-1301, USA@*
71 @node Top, Enabling OpenMP
75 This manual documents the usage of libgomp, the GNU Offloading and
76 Multi Processing Runtime Library. This includes the GNU
77 implementation of the @uref{https://www.openmp.org, OpenMP} Application
78 Programming Interface (API) for multi-platform shared-memory parallel
79 programming in C/C++ and Fortran, and the GNU implementation of the
80 @uref{https://www.openacc.org, OpenACC} Application Programming
81 Interface (API) for offloading of code to accelerator devices in C/C++
84 Originally, libgomp implemented the GNU OpenMP Runtime Library. Based
85 on this, support for OpenACC and offloading (both OpenACC and OpenMP
86 4's target construct) has been added later on, and the library's name
87 changed to GNU Offloading and Multi Processing Runtime Library.
92 @comment When you add a new menu item, please keep the right hand
93 @comment aligned to the same column. Do not use tabs. This provides
94 @comment better formatting.
97 * Enabling OpenMP:: How to enable OpenMP for your applications.
98 * OpenMP Implementation Status:: List of implemented features by OpenMP version
99 * OpenMP Runtime Library Routines: Runtime Library Routines.
100 The OpenMP runtime application programming
102 * OpenMP Environment Variables: Environment Variables.
103 Influencing OpenMP runtime behavior with
104 environment variables.
105 * Enabling OpenACC:: How to enable OpenACC for your
107 * OpenACC Runtime Library Routines:: The OpenACC runtime application
108 programming interface.
109 * OpenACC Environment Variables:: Influencing OpenACC runtime behavior with
110 environment variables.
111 * CUDA Streams Usage:: Notes on the implementation of
112 asynchronous operations.
113 * OpenACC Library Interoperability:: OpenACC library interoperability with the
114 NVIDIA CUBLAS library.
115 * OpenACC Profiling Interface::
116 * OpenMP-Implementation Specifics:: Notes specifics of this OpenMP
118 * Offload-Target Specifics:: Notes on offload-target specific internals
119 * The libgomp ABI:: Notes on the external ABI presented by libgomp.
120 * Reporting Bugs:: How to report bugs in the GNU Offloading and
121 Multi Processing Runtime Library.
122 * Copying:: GNU general public license says
123 how you can copy and share libgomp.
124 * GNU Free Documentation License::
125 How you can copy and share this manual.
126 * Funding:: How to help assure continued work for free
128 * Library Index:: Index of this documentation.
132 @c ---------------------------------------------------------------------
134 @c ---------------------------------------------------------------------
136 @node Enabling OpenMP
137 @chapter Enabling OpenMP
139 To activate the OpenMP extensions for C/C++ and Fortran, the compile-time
140 flag @option{-fopenmp} must be specified. For C and C++, this enables
141 the handling of the OpenMP directives using @code{#pragma omp} and the
142 @code{[[omp::directive(...)]]}, @code{[[omp::sequence(...)]]} and
143 @code{[[omp::decl(...)]]} attributes. For Fortran, it enables for
144 free source form the @code{!$omp} sentinel for directives and the
145 @code{!$} conditional compilation sentinel and for fixed source form the
146 @code{c$omp}, @code{*$omp} and @code{!$omp} sentinels for directives and
147 the @code{c$}, @code{*$} and @code{!$} conditional compilation sentinels.
148 The flag also arranges for automatic linking of the OpenMP runtime library
149 (@ref{Runtime Library Routines}).
151 The @option{-fopenmp-simd} flag can be used to enable a subset of
152 OpenMP directives that do not require the linking of either the
153 OpenMP runtime library or the POSIX threads library.
155 A complete description of all OpenMP directives may be found in the
156 @uref{https://www.openmp.org, OpenMP Application Program Interface} manuals.
157 See also @ref{OpenMP Implementation Status}.
160 @c ---------------------------------------------------------------------
161 @c OpenMP Implementation Status
162 @c ---------------------------------------------------------------------
164 @node OpenMP Implementation Status
165 @chapter OpenMP Implementation Status
168 * OpenMP 4.5:: Feature completion status to 4.5 specification
169 * OpenMP 5.0:: Feature completion status to 5.0 specification
170 * OpenMP 5.1:: Feature completion status to 5.1 specification
171 * OpenMP 5.2:: Feature completion status to 5.2 specification
172 * OpenMP Technical Report 12:: Feature completion status to second 6.0 preview
175 The @code{_OPENMP} preprocessor macro and Fortran's @code{openmp_version}
176 parameter, provided by @code{omp_lib.h} and the @code{omp_lib} module, have
177 the value @code{201511} (i.e. OpenMP 4.5).
182 The OpenMP 4.5 specification is fully supported.
187 @unnumberedsubsec New features listed in Appendix B of the OpenMP specification
188 @c This list is sorted as in OpenMP 5.1's B.3 not as in OpenMP 5.0's B.2
190 @multitable @columnfractions .60 .10 .25
191 @headitem Description @tab Status @tab Comments
192 @item Array shaping @tab N @tab
193 @item Array sections with non-unit strides in C and C++ @tab N @tab
194 @item Iterators @tab Y @tab
195 @item @code{metadirective} directive @tab N @tab
196 @item @code{declare variant} directive
197 @tab P @tab @emph{simd} traits not handled correctly
198 @item @var{target-offload-var} ICV and @code{OMP_TARGET_OFFLOAD}
199 env variable @tab Y @tab
200 @item Nested-parallel changes to @var{max-active-levels-var} ICV @tab Y @tab
201 @item @code{requires} directive @tab P
202 @tab complete but no non-host device provides @code{unified_shared_memory}
203 @item @code{teams} construct outside an enclosing target region @tab Y @tab
204 @item Non-rectangular loop nests @tab P
205 @tab Full support for C/C++, partial for Fortran
206 (@uref{https://gcc.gnu.org/PR110735,PR110735})
207 @item @code{!=} as relational-op in canonical loop form for C/C++ @tab Y @tab
208 @item @code{nonmonotonic} as default loop schedule modifier for worksharing-loop
209 constructs @tab Y @tab
210 @item Collapse of associated loops that are imperfectly nested loops @tab Y @tab
211 @item Clauses @code{if}, @code{nontemporal} and @code{order(concurrent)} in
212 @code{simd} construct @tab Y @tab
213 @item @code{atomic} constructs in @code{simd} @tab Y @tab
214 @item @code{loop} construct @tab Y @tab
215 @item @code{order(concurrent)} clause @tab Y @tab
216 @item @code{scan} directive and @code{in_scan} modifier for the
217 @code{reduction} clause @tab Y @tab
218 @item @code{in_reduction} clause on @code{task} constructs @tab Y @tab
219 @item @code{in_reduction} clause on @code{target} constructs @tab P
220 @tab @code{nowait} only stub
221 @item @code{task_reduction} clause with @code{taskgroup} @tab Y @tab
222 @item @code{task} modifier to @code{reduction} clause @tab Y @tab
223 @item @code{affinity} clause to @code{task} construct @tab Y @tab Stub only
224 @item @code{detach} clause to @code{task} construct @tab Y @tab
225 @item @code{omp_fulfill_event} runtime routine @tab Y @tab
226 @item @code{reduction} and @code{in_reduction} clauses on @code{taskloop}
227 and @code{taskloop simd} constructs @tab Y @tab
228 @item @code{taskloop} construct cancelable by @code{cancel} construct
230 @item @code{mutexinoutset} @emph{dependence-type} for @code{depend} clause
232 @item Predefined memory spaces, memory allocators, allocator traits
233 @tab Y @tab See also @ref{Memory allocation}
234 @item Memory management routines @tab Y @tab
235 @item @code{allocate} directive @tab P
236 @tab Only C for stack/automatic and Fortran for stack/automatic
237 and allocatable/pointer variables
238 @item @code{allocate} clause @tab P @tab Initial support
239 @item @code{use_device_addr} clause on @code{target data} @tab Y @tab
240 @item @code{ancestor} modifier on @code{device} clause @tab Y @tab
241 @item Implicit declare target directive @tab Y @tab
242 @item Discontiguous array section with @code{target update} construct
244 @item C/C++'s lvalue expressions in @code{to}, @code{from}
245 and @code{map} clauses @tab Y @tab
246 @item C/C++'s lvalue expressions in @code{depend} clauses @tab Y @tab
247 @item Nested @code{declare target} directive @tab Y @tab
248 @item Combined @code{master} constructs @tab Y @tab
249 @item @code{depend} clause on @code{taskwait} @tab Y @tab
250 @item Weak memory ordering clauses on @code{atomic} and @code{flush} construct
252 @item @code{hint} clause on the @code{atomic} construct @tab Y @tab Stub only
253 @item @code{depobj} construct and depend objects @tab Y @tab
254 @item Lock hints were renamed to synchronization hints @tab Y @tab
255 @item @code{conditional} modifier to @code{lastprivate} clause @tab Y @tab
256 @item Map-order clarifications @tab P @tab
257 @item @code{close} @emph{map-type-modifier} @tab Y @tab
258 @item Mapping C/C++ pointer variables and to assign the address of
259 device memory mapped by an array section @tab P @tab
260 @item Mapping of Fortran pointer and allocatable variables, including pointer
261 and allocatable components of variables
262 @tab P @tab Mapping of vars with allocatable components unsupported
263 @item @code{defaultmap} extensions @tab Y @tab
264 @item @code{declare mapper} directive @tab N @tab
265 @item @code{omp_get_supported_active_levels} routine @tab Y @tab
266 @item Runtime routines and environment variables to display runtime thread
267 affinity information @tab Y @tab
268 @item @code{omp_pause_resource} and @code{omp_pause_resource_all} runtime
270 @item @code{omp_get_device_num} runtime routine @tab Y @tab
271 @item OMPT interface @tab N @tab
272 @item OMPD interface @tab N @tab
275 @unnumberedsubsec Other new OpenMP 5.0 features
277 @multitable @columnfractions .60 .10 .25
278 @headitem Description @tab Status @tab Comments
279 @item Supporting C++'s range-based for loop @tab Y @tab
286 @unnumberedsubsec New features listed in Appendix B of the OpenMP specification
288 @multitable @columnfractions .60 .10 .25
289 @headitem Description @tab Status @tab Comments
290 @item OpenMP directive as C++ attribute specifiers @tab Y @tab
291 @item @code{omp_all_memory} reserved locator @tab Y @tab
292 @item @emph{target_device trait} in OpenMP Context @tab N @tab
293 @item @code{target_device} selector set in context selectors @tab N @tab
294 @item C/C++'s @code{declare variant} directive: elision support of
295 preprocessed code @tab N @tab
296 @item @code{declare variant}: new clauses @code{adjust_args} and
297 @code{append_args} @tab N @tab
298 @item @code{dispatch} construct @tab N @tab
299 @item device-specific ICV settings with environment variables @tab Y @tab
300 @item @code{assume} and @code{assumes} directives @tab Y @tab
301 @item @code{nothing} directive @tab Y @tab
302 @item @code{error} directive @tab Y @tab
303 @item @code{masked} construct @tab Y @tab
304 @item @code{scope} directive @tab Y @tab
305 @item Loop transformation constructs @tab N @tab
306 @item @code{strict} modifier in the @code{grainsize} and @code{num_tasks}
307 clauses of the @code{taskloop} construct @tab Y @tab
308 @item @code{align} clause in @code{allocate} directive @tab P
309 @tab Only C and Fortran (and not for static variables)
310 @item @code{align} modifier in @code{allocate} clause @tab Y @tab
311 @item @code{thread_limit} clause to @code{target} construct @tab Y @tab
312 @item @code{has_device_addr} clause to @code{target} construct @tab Y @tab
313 @item Iterators in @code{target update} motion clauses and @code{map}
315 @item Indirect calls to the device version of a procedure or function in
316 @code{target} regions @tab P @tab Only C and C++
317 @item @code{interop} directive @tab N @tab
318 @item @code{omp_interop_t} object support in runtime routines @tab N @tab
319 @item @code{nowait} clause in @code{taskwait} directive @tab Y @tab
320 @item Extensions to the @code{atomic} directive @tab Y @tab
321 @item @code{seq_cst} clause on a @code{flush} construct @tab Y @tab
322 @item @code{inoutset} argument to the @code{depend} clause @tab Y @tab
323 @item @code{private} and @code{firstprivate} argument to @code{default}
324 clause in C and C++ @tab Y @tab
325 @item @code{present} argument to @code{defaultmap} clause @tab Y @tab
326 @item @code{omp_set_num_teams}, @code{omp_set_teams_thread_limit},
327 @code{omp_get_max_teams}, @code{omp_get_teams_thread_limit} runtime
329 @item @code{omp_target_is_accessible} runtime routine @tab Y @tab
330 @item @code{omp_target_memcpy_async} and @code{omp_target_memcpy_rect_async}
331 runtime routines @tab Y @tab
332 @item @code{omp_get_mapped_ptr} runtime routine @tab Y @tab
333 @item @code{omp_calloc}, @code{omp_realloc}, @code{omp_aligned_alloc} and
334 @code{omp_aligned_calloc} runtime routines @tab Y @tab
335 @item @code{omp_alloctrait_key_t} enum: @code{omp_atv_serialized} added,
336 @code{omp_atv_default} changed @tab Y @tab
337 @item @code{omp_display_env} runtime routine @tab Y @tab
338 @item @code{ompt_scope_endpoint_t} enum: @code{ompt_scope_beginend} @tab N @tab
339 @item @code{ompt_sync_region_t} enum additions @tab N @tab
340 @item @code{ompt_state_t} enum: @code{ompt_state_wait_barrier_implementation}
341 and @code{ompt_state_wait_barrier_teams} @tab N @tab
342 @item @code{ompt_callback_target_data_op_emi_t},
343 @code{ompt_callback_target_emi_t}, @code{ompt_callback_target_map_emi_t}
344 and @code{ompt_callback_target_submit_emi_t} @tab N @tab
345 @item @code{ompt_callback_error_t} type @tab N @tab
346 @item @code{OMP_PLACES} syntax extensions @tab Y @tab
347 @item @code{OMP_NUM_TEAMS} and @code{OMP_TEAMS_THREAD_LIMIT} environment
348 variables @tab Y @tab
351 @unnumberedsubsec Other new OpenMP 5.1 features
353 @multitable @columnfractions .60 .10 .25
354 @headitem Description @tab Status @tab Comments
355 @item Support of strictly structured blocks in Fortran @tab Y @tab
356 @item Support of structured block sequences in C/C++ @tab Y @tab
357 @item @code{unconstrained} and @code{reproducible} modifiers on @code{order}
359 @item Support @code{begin/end declare target} syntax in C/C++ @tab Y @tab
360 @item Pointer predetermined firstprivate getting initialized
361 to address of matching mapped list item per 5.1, Sect. 2.21.7.2 @tab N @tab
362 @item For Fortran, diagnose placing declarative before/between @code{USE},
363 @code{IMPORT}, and @code{IMPLICIT} as invalid @tab N @tab
364 @item Optional comma between directive and clause in the @code{#pragma} form @tab Y @tab
365 @item @code{indirect} clause in @code{declare target} @tab P @tab Only C and C++
366 @item @code{device_type(nohost)}/@code{device_type(host)} for variables @tab N @tab
367 @item @code{present} modifier to the @code{map}, @code{to} and @code{from}
375 @unnumberedsubsec New features listed in Appendix B of the OpenMP specification
377 @multitable @columnfractions .60 .10 .25
378 @headitem Description @tab Status @tab Comments
379 @item @code{omp_in_explicit_task} routine and @var{explicit-task-var} ICV
381 @item @code{omp}/@code{ompx}/@code{omx} sentinels and @code{omp_}/@code{ompx_}
383 @tab warning for @code{ompx/omx} sentinels@footnote{The @code{ompx}
384 sentinel as C/C++ pragma and C++ attributes are warned for with
385 @code{-Wunknown-pragmas} (implied by @code{-Wall}) and @code{-Wattributes}
386 (enabled by default), respectively; for Fortran free-source code, there is
387 a warning enabled by default and, for fixed-source code, the @code{omx}
388 sentinel is warned for with with @code{-Wsurprising} (enabled by
389 @code{-Wall}). Unknown clauses are always rejected with an error.}
390 @item Clauses on @code{end} directive can be on directive @tab Y @tab
391 @item @code{destroy} clause with destroy-var argument on @code{depobj}
393 @item Deprecation of no-argument @code{destroy} clause on @code{depobj}
395 @item @code{linear} clause syntax changes and @code{step} modifier @tab Y @tab
396 @item Deprecation of minus operator for reductions @tab N @tab
397 @item Deprecation of separating @code{map} modifiers without comma @tab N @tab
398 @item @code{declare mapper} with iterator and @code{present} modifiers
400 @item If a matching mapped list item is not found in the data environment, the
401 pointer retains its original value @tab Y @tab
402 @item New @code{enter} clause as alias for @code{to} on declare target directive
404 @item Deprecation of @code{to} clause on declare target directive @tab N @tab
405 @item Extended list of directives permitted in Fortran pure procedures
407 @item New @code{allocators} directive for Fortran @tab Y @tab
408 @item Deprecation of @code{allocate} directive for Fortran
409 allocatables/pointers @tab N @tab
410 @item Optional paired @code{end} directive with @code{dispatch} @tab N @tab
411 @item New @code{memspace} and @code{traits} modifiers for @code{uses_allocators}
413 @item Deprecation of traits array following the allocator_handle expression in
414 @code{uses_allocators} @tab N @tab
415 @item New @code{otherwise} clause as alias for @code{default} on metadirectives
417 @item Deprecation of @code{default} clause on metadirectives @tab N @tab
418 @item Deprecation of delimited form of @code{declare target} @tab N @tab
419 @item Reproducible semantics changed for @code{order(concurrent)} @tab N @tab
420 @item @code{allocate} and @code{firstprivate} clauses on @code{scope}
422 @item @code{ompt_callback_work} @tab N @tab
423 @item Default map-type for the @code{map} clause in @code{target enter/exit data}
425 @item New @code{doacross} clause as alias for @code{depend} with
426 @code{source}/@code{sink} modifier @tab Y @tab
427 @item Deprecation of @code{depend} with @code{source}/@code{sink} modifier
429 @item @code{omp_cur_iteration} keyword @tab Y @tab
432 @unnumberedsubsec Other new OpenMP 5.2 features
434 @multitable @columnfractions .60 .10 .25
435 @headitem Description @tab Status @tab Comments
436 @item For Fortran, optional comma between directive and clause @tab N @tab
437 @item Conforming device numbers and @code{omp_initial_device} and
438 @code{omp_invalid_device} enum/PARAMETER @tab Y @tab
439 @item Initial value of @var{default-device-var} ICV with
440 @code{OMP_TARGET_OFFLOAD=mandatory} @tab Y @tab
441 @item @code{all} as @emph{implicit-behavior} for @code{defaultmap} @tab Y @tab
442 @item @emph{interop_types} in any position of the modifier list for the @code{init} clause
443 of the @code{interop} construct @tab N @tab
444 @item Invoke virtual member functions of C++ objects created on the host device
445 on other devices @tab N @tab
449 @node OpenMP Technical Report 12
450 @section OpenMP Technical Report 12
452 Technical Report (TR) 12 is the second preview for OpenMP 6.0.
454 @unnumberedsubsec New features listed in Appendix B of the OpenMP specification
455 @multitable @columnfractions .60 .10 .25
456 @item Features deprecated in versions 5.2, 5.1 and 5.0 were removed
457 @tab N/A @tab Backward compatibility
458 @item Full support for C23 was added @tab P @tab
459 @item Full support for C++23 was added @tab P @tab
460 @item @code{_ALL} suffix to the device-scope environment variables
461 @tab P @tab Host device number wrongly accepted
462 @item @code{num_threads} now accepts a list @tab N @tab
463 @item Supporting increments with abstract names in @code{OMP_PLACES} @tab N @tab
464 @item Extension of @code{OMP_DEFAULT_DEVICE} and new
465 @code{OMP_AVAILABLE_DEVICES} environment vars @tab N @tab
466 @item New @code{OMP_THREADS_RESERVE} environment variable @tab N @tab
467 @item The @code{decl} attribute was added to the C++ attribute syntax
469 @item The OpenMP directive syntax was extended to include C 23 attribute
470 specifiers @tab Y @tab
471 @item All inarguable clauses take now an optional Boolean argument @tab N @tab
472 @item For Fortran, @emph{locator list} can be also function reference with
473 data pointer result @tab N @tab
474 @item Concept of @emph{assumed-size arrays} in C and C++
476 @item @emph{directive-name-modifier} accepted in all clauses @tab N @tab
477 @item For Fortran, atomic with BLOCK construct and, for C/C++, with
478 unlimited curly braces supported @tab N @tab
479 @item For Fortran, atomic compare with storing the comparison result
481 @item New @code{looprange} clause @tab N @tab
482 @item Ref-count change for @code{use_device_ptr}/@code{use_device_addr}
484 @item Support for inductions @tab N @tab
485 @item Implicit reduction identifiers of C++ classes
487 @item Change of the @emph{map-type} property from @emph{ultimate} to
488 @emph{default} @tab N @tab
489 @item @code{self} modifier to @code{map} and @code{self} as
490 @code{defaultmap} argument @tab N @tab
491 @item Mapping of @emph{assumed-size arrays} in C, C++ and Fortran
493 @item @code{groupprivate} directive @tab N @tab
494 @item @code{local} clause to @code{declare target} directive @tab N @tab
495 @item @code{part_size} allocator trait @tab N @tab
496 @item @code{pin_device}, @code{preferred_device} and @code{target_access}
499 @item @code{access} allocator trait changes @tab N @tab
500 @item Extension of @code{interop} operation of @code{append_args}, allowing all
501 modifiers of the @code{init} clause
503 @item @code{interop} clause to @code{dispatch} @tab N @tab
504 @item @code{message} and @code{severity} clauses to @code{parallel} directive
506 @item @code{self} clause to @code{requires} directive @tab N @tab
507 @item @code{no_openmp_constructs} assumptions clause @tab N @tab
508 @item @code{reverse} loop-transformation construct @tab N @tab
509 @item @code{interchange} loop-transformation construct @tab N @tab
510 @item @code{fuse} loop-transformation construct @tab N @tab
511 @item @code{apply} code to loop-transforming constructs @tab N @tab
512 @item @code{omp_curr_progress_width} identifier @tab N @tab
513 @item @code{safesync} clause to the @code{parallel} construct @tab N @tab
514 @item @code{omp_get_max_progress_width} runtime routine @tab N @tab
515 @item @code{strict} modifier keyword to @code{num_threads} @tab N @tab
516 @item @code{atomic} permitted in a construct with @code{order(concurrent)}
518 @item @code{coexecute} directive for Fortran @tab N @tab
519 @item Fortran DO CONCURRENT as associated loop in a @code{loop} construct
521 @item @code{threadset} clause in task-generating constructs @tab N @tab
522 @item @code{nowait} clause with reverse-offload @code{target} directives
524 @item Boolean argument to @code{nowait} and @code{nogroup} may be non constant
526 @item @code{memscope} clause to @code{atomic} and @code{flush} @tab N @tab
527 @item @code{omp_is_free_agent} and @code{omp_ancestor_is_free_agent} routines
529 @item @code{omp_target_memset} and @code{omp_target_memset_rect_async} routines
531 @item Routines for obtaining memory spaces/allocators for shared/device memory
533 @item @code{omp_get_memspace_num_resources} routine @tab N @tab
534 @item @code{omp_get_submemspace} routine @tab N @tab
535 @item @code{ompt_target_data_transfer} and @code{ompt_target_data_transfer_async}
536 values in @code{ompt_target_data_op_t} enum @tab N @tab
537 @item @code{ompt_get_buffer_limits} OMPT routine @tab N @tab
540 @unnumberedsubsec Other new TR 12 features
541 @multitable @columnfractions .60 .10 .25
542 @item Relaxed Fortran restrictions to the @code{aligned} clause @tab N @tab
543 @item Mapping lambda captures @tab N @tab
544 @item New @code{omp_pause_stop_tool} constant for omp_pause_resource @tab N @tab
549 @c ---------------------------------------------------------------------
550 @c OpenMP Runtime Library Routines
551 @c ---------------------------------------------------------------------
553 @node Runtime Library Routines
554 @chapter OpenMP Runtime Library Routines
556 The runtime routines described here are defined by Section 18 of the OpenMP
557 specification in version 5.2.
560 * Thread Team Routines::
561 * Thread Affinity Routines::
562 * Teams Region Routines::
564 * Resource Relinquishing Routines::
565 * Device Information Routines::
566 * Device Memory Routines::
570 @c * Interoperability Routines::
571 * Memory Management Routines::
572 @c * Tool Control Routine::
573 * Environment Display Routine::
578 @node Thread Team Routines
579 @section Thread Team Routines
581 Routines controlling threads in the current contention group.
582 They have C linkage and do not throw exceptions.
585 * omp_set_num_threads:: Set upper team size limit
586 * omp_get_num_threads:: Size of the active team
587 * omp_get_max_threads:: Maximum number of threads of parallel region
588 * omp_get_thread_num:: Current thread ID
589 * omp_in_parallel:: Whether a parallel region is active
590 * omp_set_dynamic:: Enable/disable dynamic teams
591 * omp_get_dynamic:: Dynamic teams setting
592 * omp_get_cancellation:: Whether cancellation support is enabled
593 * omp_set_nested:: Enable/disable nested parallel regions
594 * omp_get_nested:: Nested parallel regions
595 * omp_set_schedule:: Set the runtime scheduling method
596 * omp_get_schedule:: Obtain the runtime scheduling method
597 * omp_get_teams_thread_limit:: Maximum number of threads imposed by teams
598 * omp_get_supported_active_levels:: Maximum number of active regions supported
599 * omp_set_max_active_levels:: Limits the number of active parallel regions
600 * omp_get_max_active_levels:: Current maximum number of active regions
601 * omp_get_level:: Number of parallel regions
602 * omp_get_ancestor_thread_num:: Ancestor thread ID
603 * omp_get_team_size:: Number of threads in a team
604 * omp_get_active_level:: Number of active parallel regions
609 @node omp_set_num_threads
610 @subsection @code{omp_set_num_threads} -- Set upper team size limit
612 @item @emph{Description}:
613 Specifies the number of threads used by default in subsequent parallel
614 sections, if those do not specify a @code{num_threads} clause. The
615 argument of @code{omp_set_num_threads} shall be a positive integer.
618 @multitable @columnfractions .20 .80
619 @item @emph{Prototype}: @tab @code{void omp_set_num_threads(int num_threads);}
622 @item @emph{Fortran}:
623 @multitable @columnfractions .20 .80
624 @item @emph{Interface}: @tab @code{subroutine omp_set_num_threads(num_threads)}
625 @item @tab @code{integer, intent(in) :: num_threads}
628 @item @emph{See also}:
629 @ref{OMP_NUM_THREADS}, @ref{omp_get_num_threads}, @ref{omp_get_max_threads}
631 @item @emph{Reference}:
632 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.1.
637 @node omp_get_num_threads
638 @subsection @code{omp_get_num_threads} -- Size of the active team
640 @item @emph{Description}:
641 Returns the number of threads in the current team. In a sequential section of
642 the program @code{omp_get_num_threads} returns 1.
644 The default team size may be initialized at startup by the
645 @env{OMP_NUM_THREADS} environment variable. At runtime, the size
646 of the current team may be set either by the @code{NUM_THREADS}
647 clause or by @code{omp_set_num_threads}. If none of the above were
648 used to define a specific value and @env{OMP_DYNAMIC} is disabled,
649 one thread per CPU online is used.
652 @multitable @columnfractions .20 .80
653 @item @emph{Prototype}: @tab @code{int omp_get_num_threads(void);}
656 @item @emph{Fortran}:
657 @multitable @columnfractions .20 .80
658 @item @emph{Interface}: @tab @code{integer function omp_get_num_threads()}
661 @item @emph{See also}:
662 @ref{omp_get_max_threads}, @ref{omp_set_num_threads}, @ref{OMP_NUM_THREADS}
664 @item @emph{Reference}:
665 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.2.
670 @node omp_get_max_threads
671 @subsection @code{omp_get_max_threads} -- Maximum number of threads of parallel region
673 @item @emph{Description}:
674 Return the maximum number of threads used for the current parallel region
675 that does not use the clause @code{num_threads}.
678 @multitable @columnfractions .20 .80
679 @item @emph{Prototype}: @tab @code{int omp_get_max_threads(void);}
682 @item @emph{Fortran}:
683 @multitable @columnfractions .20 .80
684 @item @emph{Interface}: @tab @code{integer function omp_get_max_threads()}
687 @item @emph{See also}:
688 @ref{omp_set_num_threads}, @ref{omp_set_dynamic}, @ref{omp_get_thread_limit}
690 @item @emph{Reference}:
691 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.3.
696 @node omp_get_thread_num
697 @subsection @code{omp_get_thread_num} -- Current thread ID
699 @item @emph{Description}:
700 Returns a unique thread identification number within the current team.
701 In a sequential parts of the program, @code{omp_get_thread_num}
702 always returns 0. In parallel regions the return value varies
703 from 0 to @code{omp_get_num_threads}-1 inclusive. The return
704 value of the primary thread of a team is always 0.
707 @multitable @columnfractions .20 .80
708 @item @emph{Prototype}: @tab @code{int omp_get_thread_num(void);}
711 @item @emph{Fortran}:
712 @multitable @columnfractions .20 .80
713 @item @emph{Interface}: @tab @code{integer function omp_get_thread_num()}
716 @item @emph{See also}:
717 @ref{omp_get_num_threads}, @ref{omp_get_ancestor_thread_num}
719 @item @emph{Reference}:
720 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.4.
725 @node omp_in_parallel
726 @subsection @code{omp_in_parallel} -- Whether a parallel region is active
728 @item @emph{Description}:
729 This function returns @code{true} if currently running in parallel,
730 @code{false} otherwise. Here, @code{true} and @code{false} represent
731 their language-specific counterparts.
734 @multitable @columnfractions .20 .80
735 @item @emph{Prototype}: @tab @code{int omp_in_parallel(void);}
738 @item @emph{Fortran}:
739 @multitable @columnfractions .20 .80
740 @item @emph{Interface}: @tab @code{logical function omp_in_parallel()}
743 @item @emph{Reference}:
744 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.6.
748 @node omp_set_dynamic
749 @subsection @code{omp_set_dynamic} -- Enable/disable dynamic teams
751 @item @emph{Description}:
752 Enable or disable the dynamic adjustment of the number of threads
753 within a team. The function takes the language-specific equivalent
754 of @code{true} and @code{false}, where @code{true} enables dynamic
755 adjustment of team sizes and @code{false} disables it.
758 @multitable @columnfractions .20 .80
759 @item @emph{Prototype}: @tab @code{void omp_set_dynamic(int dynamic_threads);}
762 @item @emph{Fortran}:
763 @multitable @columnfractions .20 .80
764 @item @emph{Interface}: @tab @code{subroutine omp_set_dynamic(dynamic_threads)}
765 @item @tab @code{logical, intent(in) :: dynamic_threads}
768 @item @emph{See also}:
769 @ref{OMP_DYNAMIC}, @ref{omp_get_dynamic}
771 @item @emph{Reference}:
772 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.7.
777 @node omp_get_dynamic
778 @subsection @code{omp_get_dynamic} -- Dynamic teams setting
780 @item @emph{Description}:
781 This function returns @code{true} if enabled, @code{false} otherwise.
782 Here, @code{true} and @code{false} represent their language-specific
785 The dynamic team setting may be initialized at startup by the
786 @env{OMP_DYNAMIC} environment variable or at runtime using
787 @code{omp_set_dynamic}. If undefined, dynamic adjustment is
791 @multitable @columnfractions .20 .80
792 @item @emph{Prototype}: @tab @code{int omp_get_dynamic(void);}
795 @item @emph{Fortran}:
796 @multitable @columnfractions .20 .80
797 @item @emph{Interface}: @tab @code{logical function omp_get_dynamic()}
800 @item @emph{See also}:
801 @ref{omp_set_dynamic}, @ref{OMP_DYNAMIC}
803 @item @emph{Reference}:
804 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.8.
809 @node omp_get_cancellation
810 @subsection @code{omp_get_cancellation} -- Whether cancellation support is enabled
812 @item @emph{Description}:
813 This function returns @code{true} if cancellation is activated, @code{false}
814 otherwise. Here, @code{true} and @code{false} represent their language-specific
815 counterparts. Unless @env{OMP_CANCELLATION} is set true, cancellations are
819 @multitable @columnfractions .20 .80
820 @item @emph{Prototype}: @tab @code{int omp_get_cancellation(void);}
823 @item @emph{Fortran}:
824 @multitable @columnfractions .20 .80
825 @item @emph{Interface}: @tab @code{logical function omp_get_cancellation()}
828 @item @emph{See also}:
829 @ref{OMP_CANCELLATION}
831 @item @emph{Reference}:
832 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.9.
838 @subsection @code{omp_set_nested} -- Enable/disable nested parallel regions
840 @item @emph{Description}:
841 Enable or disable nested parallel regions, i.e., whether team members
842 are allowed to create new teams. The function takes the language-specific
843 equivalent of @code{true} and @code{false}, where @code{true} enables
844 dynamic adjustment of team sizes and @code{false} disables it.
846 Enabling nested parallel regions also sets the maximum number of
847 active nested regions to the maximum supported. Disabling nested parallel
848 regions sets the maximum number of active nested regions to one.
850 Note that the @code{omp_set_nested} API routine was deprecated
851 in the OpenMP specification 5.2 in favor of @code{omp_set_max_active_levels}.
854 @multitable @columnfractions .20 .80
855 @item @emph{Prototype}: @tab @code{void omp_set_nested(int nested);}
858 @item @emph{Fortran}:
859 @multitable @columnfractions .20 .80
860 @item @emph{Interface}: @tab @code{subroutine omp_set_nested(nested)}
861 @item @tab @code{logical, intent(in) :: nested}
864 @item @emph{See also}:
865 @ref{omp_get_nested}, @ref{omp_set_max_active_levels},
866 @ref{OMP_MAX_ACTIVE_LEVELS}, @ref{OMP_NESTED}
868 @item @emph{Reference}:
869 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.10.
875 @subsection @code{omp_get_nested} -- Nested parallel regions
877 @item @emph{Description}:
878 This function returns @code{true} if nested parallel regions are
879 enabled, @code{false} otherwise. Here, @code{true} and @code{false}
880 represent their language-specific counterparts.
882 The state of nested parallel regions at startup depends on several
883 environment variables. If @env{OMP_MAX_ACTIVE_LEVELS} is defined
884 and is set to greater than one, then nested parallel regions will be
885 enabled. If not defined, then the value of the @env{OMP_NESTED}
886 environment variable will be followed if defined. If neither are
887 defined, then if either @env{OMP_NUM_THREADS} or @env{OMP_PROC_BIND}
888 are defined with a list of more than one value, then nested parallel
889 regions are enabled. If none of these are defined, then nested parallel
890 regions are disabled by default.
892 Nested parallel regions can be enabled or disabled at runtime using
893 @code{omp_set_nested}, or by setting the maximum number of nested
894 regions with @code{omp_set_max_active_levels} to one to disable, or
897 Note that the @code{omp_get_nested} API routine was deprecated
898 in the OpenMP specification 5.2 in favor of @code{omp_get_max_active_levels}.
901 @multitable @columnfractions .20 .80
902 @item @emph{Prototype}: @tab @code{int omp_get_nested(void);}
905 @item @emph{Fortran}:
906 @multitable @columnfractions .20 .80
907 @item @emph{Interface}: @tab @code{logical function omp_get_nested()}
910 @item @emph{See also}:
911 @ref{omp_get_max_active_levels}, @ref{omp_set_nested},
912 @ref{OMP_MAX_ACTIVE_LEVELS}, @ref{OMP_NESTED}
914 @item @emph{Reference}:
915 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.11.
920 @node omp_set_schedule
921 @subsection @code{omp_set_schedule} -- Set the runtime scheduling method
923 @item @emph{Description}:
924 Sets the runtime scheduling method. The @var{kind} argument can have the
925 value @code{omp_sched_static}, @code{omp_sched_dynamic},
926 @code{omp_sched_guided} or @code{omp_sched_auto}. Except for
927 @code{omp_sched_auto}, the chunk size is set to the value of
928 @var{chunk_size} if positive, or to the default value if zero or negative.
929 For @code{omp_sched_auto} the @var{chunk_size} argument is ignored.
932 @multitable @columnfractions .20 .80
933 @item @emph{Prototype}: @tab @code{void omp_set_schedule(omp_sched_t kind, int chunk_size);}
936 @item @emph{Fortran}:
937 @multitable @columnfractions .20 .80
938 @item @emph{Interface}: @tab @code{subroutine omp_set_schedule(kind, chunk_size)}
939 @item @tab @code{integer(kind=omp_sched_kind) kind}
940 @item @tab @code{integer chunk_size}
943 @item @emph{See also}:
944 @ref{omp_get_schedule}
947 @item @emph{Reference}:
948 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.12.
953 @node omp_get_schedule
954 @subsection @code{omp_get_schedule} -- Obtain the runtime scheduling method
956 @item @emph{Description}:
957 Obtain the runtime scheduling method. The @var{kind} argument is set to
958 @code{omp_sched_static}, @code{omp_sched_dynamic},
959 @code{omp_sched_guided} or @code{omp_sched_auto}. The second argument,
960 @var{chunk_size}, is set to the chunk size.
963 @multitable @columnfractions .20 .80
964 @item @emph{Prototype}: @tab @code{void omp_get_schedule(omp_sched_t *kind, int *chunk_size);}
967 @item @emph{Fortran}:
968 @multitable @columnfractions .20 .80
969 @item @emph{Interface}: @tab @code{subroutine omp_get_schedule(kind, chunk_size)}
970 @item @tab @code{integer(kind=omp_sched_kind) kind}
971 @item @tab @code{integer chunk_size}
974 @item @emph{See also}:
975 @ref{omp_set_schedule}, @ref{OMP_SCHEDULE}
977 @item @emph{Reference}:
978 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.13.
982 @node omp_get_teams_thread_limit
983 @subsection @code{omp_get_teams_thread_limit} -- Maximum number of threads imposed by teams
985 @item @emph{Description}:
986 Return the maximum number of threads that are able to participate in
987 each team created by a teams construct.
990 @multitable @columnfractions .20 .80
991 @item @emph{Prototype}: @tab @code{int omp_get_teams_thread_limit(void);}
994 @item @emph{Fortran}:
995 @multitable @columnfractions .20 .80
996 @item @emph{Interface}: @tab @code{integer function omp_get_teams_thread_limit()}
999 @item @emph{See also}:
1000 @ref{omp_set_teams_thread_limit}, @ref{OMP_TEAMS_THREAD_LIMIT}
1002 @item @emph{Reference}:
1003 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.4.6.
1008 @node omp_get_supported_active_levels
1009 @subsection @code{omp_get_supported_active_levels} -- Maximum number of active regions supported
1011 @item @emph{Description}:
1012 This function returns the maximum number of nested, active parallel regions
1013 supported by this implementation.
1016 @multitable @columnfractions .20 .80
1017 @item @emph{Prototype}: @tab @code{int omp_get_supported_active_levels(void);}
1020 @item @emph{Fortran}:
1021 @multitable @columnfractions .20 .80
1022 @item @emph{Interface}: @tab @code{integer function omp_get_supported_active_levels()}
1025 @item @emph{See also}:
1026 @ref{omp_get_max_active_levels}, @ref{omp_set_max_active_levels}
1028 @item @emph{Reference}:
1029 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.2.15.
1034 @node omp_set_max_active_levels
1035 @subsection @code{omp_set_max_active_levels} -- Limits the number of active parallel regions
1037 @item @emph{Description}:
1038 This function limits the maximum allowed number of nested, active
1039 parallel regions. @var{max_levels} must be less or equal to
1040 the value returned by @code{omp_get_supported_active_levels}.
1043 @multitable @columnfractions .20 .80
1044 @item @emph{Prototype}: @tab @code{void omp_set_max_active_levels(int max_levels);}
1047 @item @emph{Fortran}:
1048 @multitable @columnfractions .20 .80
1049 @item @emph{Interface}: @tab @code{subroutine omp_set_max_active_levels(max_levels)}
1050 @item @tab @code{integer max_levels}
1053 @item @emph{See also}:
1054 @ref{omp_get_max_active_levels}, @ref{omp_get_active_level},
1055 @ref{omp_get_supported_active_levels}
1057 @item @emph{Reference}:
1058 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.15.
1063 @node omp_get_max_active_levels
1064 @subsection @code{omp_get_max_active_levels} -- Current maximum number of active regions
1066 @item @emph{Description}:
1067 This function obtains the maximum allowed number of nested, active parallel regions.
1070 @multitable @columnfractions .20 .80
1071 @item @emph{Prototype}: @tab @code{int omp_get_max_active_levels(void);}
1074 @item @emph{Fortran}:
1075 @multitable @columnfractions .20 .80
1076 @item @emph{Interface}: @tab @code{integer function omp_get_max_active_levels()}
1079 @item @emph{See also}:
1080 @ref{omp_set_max_active_levels}, @ref{omp_get_active_level}
1082 @item @emph{Reference}:
1083 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.16.
1088 @subsection @code{omp_get_level} -- Obtain the current nesting level
1090 @item @emph{Description}:
1091 This function returns the nesting level for the parallel blocks,
1092 which enclose the calling call.
1095 @multitable @columnfractions .20 .80
1096 @item @emph{Prototype}: @tab @code{int omp_get_level(void);}
1099 @item @emph{Fortran}:
1100 @multitable @columnfractions .20 .80
1101 @item @emph{Interface}: @tab @code{integer function omp_level()}
1104 @item @emph{See also}:
1105 @ref{omp_get_active_level}
1107 @item @emph{Reference}:
1108 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.17.
1113 @node omp_get_ancestor_thread_num
1114 @subsection @code{omp_get_ancestor_thread_num} -- Ancestor thread ID
1116 @item @emph{Description}:
1117 This function returns the thread identification number for the given
1118 nesting level of the current thread. For values of @var{level} outside
1119 zero to @code{omp_get_level} -1 is returned; if @var{level} is
1120 @code{omp_get_level} the result is identical to @code{omp_get_thread_num}.
1123 @multitable @columnfractions .20 .80
1124 @item @emph{Prototype}: @tab @code{int omp_get_ancestor_thread_num(int level);}
1127 @item @emph{Fortran}:
1128 @multitable @columnfractions .20 .80
1129 @item @emph{Interface}: @tab @code{integer function omp_get_ancestor_thread_num(level)}
1130 @item @tab @code{integer level}
1133 @item @emph{See also}:
1134 @ref{omp_get_level}, @ref{omp_get_thread_num}, @ref{omp_get_team_size}
1136 @item @emph{Reference}:
1137 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.18.
1142 @node omp_get_team_size
1143 @subsection @code{omp_get_team_size} -- Number of threads in a team
1145 @item @emph{Description}:
1146 This function returns the number of threads in a thread team to which
1147 either the current thread or its ancestor belongs. For values of @var{level}
1148 outside zero to @code{omp_get_level}, -1 is returned; if @var{level} is zero,
1149 1 is returned, and for @code{omp_get_level}, the result is identical
1150 to @code{omp_get_num_threads}.
1153 @multitable @columnfractions .20 .80
1154 @item @emph{Prototype}: @tab @code{int omp_get_team_size(int level);}
1157 @item @emph{Fortran}:
1158 @multitable @columnfractions .20 .80
1159 @item @emph{Interface}: @tab @code{integer function omp_get_team_size(level)}
1160 @item @tab @code{integer level}
1163 @item @emph{See also}:
1164 @ref{omp_get_num_threads}, @ref{omp_get_level}, @ref{omp_get_ancestor_thread_num}
1166 @item @emph{Reference}:
1167 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.19.
1172 @node omp_get_active_level
1173 @subsection @code{omp_get_active_level} -- Number of parallel regions
1175 @item @emph{Description}:
1176 This function returns the nesting level for the active parallel blocks,
1177 which enclose the calling call.
1180 @multitable @columnfractions .20 .80
1181 @item @emph{Prototype}: @tab @code{int omp_get_active_level(void);}
1184 @item @emph{Fortran}:
1185 @multitable @columnfractions .20 .80
1186 @item @emph{Interface}: @tab @code{integer function omp_get_active_level()}
1189 @item @emph{See also}:
1190 @ref{omp_get_level}, @ref{omp_get_max_active_levels}, @ref{omp_set_max_active_levels}
1192 @item @emph{Reference}:
1193 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.20.
1198 @node Thread Affinity Routines
1199 @section Thread Affinity Routines
1201 Routines controlling and accessing thread-affinity policies.
1202 They have C linkage and do not throw exceptions.
1205 * omp_get_proc_bind:: Whether threads may be moved between CPUs
1206 @c * omp_get_num_places:: <fixme>
1207 @c * omp_get_place_num_procs:: <fixme>
1208 @c * omp_get_place_proc_ids:: <fixme>
1209 @c * omp_get_place_num:: <fixme>
1210 @c * omp_get_partition_num_places:: <fixme>
1211 @c * omp_get_partition_place_nums:: <fixme>
1212 @c * omp_set_affinity_format:: <fixme>
1213 @c * omp_get_affinity_format:: <fixme>
1214 @c * omp_display_affinity:: <fixme>
1215 @c * omp_capture_affinity:: <fixme>
1220 @node omp_get_proc_bind
1221 @subsection @code{omp_get_proc_bind} -- Whether threads may be moved between CPUs
1223 @item @emph{Description}:
1224 This functions returns the currently active thread affinity policy, which is
1225 set via @env{OMP_PROC_BIND}. Possible values are @code{omp_proc_bind_false},
1226 @code{omp_proc_bind_true}, @code{omp_proc_bind_primary},
1227 @code{omp_proc_bind_master}, @code{omp_proc_bind_close} and @code{omp_proc_bind_spread},
1228 where @code{omp_proc_bind_master} is an alias for @code{omp_proc_bind_primary}.
1231 @multitable @columnfractions .20 .80
1232 @item @emph{Prototype}: @tab @code{omp_proc_bind_t omp_get_proc_bind(void);}
1235 @item @emph{Fortran}:
1236 @multitable @columnfractions .20 .80
1237 @item @emph{Interface}: @tab @code{integer(kind=omp_proc_bind_kind) function omp_get_proc_bind()}
1240 @item @emph{See also}:
1241 @ref{OMP_PROC_BIND}, @ref{OMP_PLACES}, @ref{GOMP_CPU_AFFINITY},
1243 @item @emph{Reference}:
1244 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.22.
1249 @node Teams Region Routines
1250 @section Teams Region Routines
1252 Routines controlling the league of teams that are executed in a @code{teams}
1253 region. They have C linkage and do not throw exceptions.
1256 * omp_get_num_teams:: Number of teams
1257 * omp_get_team_num:: Get team number
1258 * omp_set_num_teams:: Set upper teams limit for teams region
1259 * omp_get_max_teams:: Maximum number of teams for teams region
1260 * omp_set_teams_thread_limit:: Set upper thread limit for teams construct
1261 * omp_get_thread_limit:: Maximum number of threads
1266 @node omp_get_num_teams
1267 @subsection @code{omp_get_num_teams} -- Number of teams
1269 @item @emph{Description}:
1270 Returns the number of teams in the current team region.
1273 @multitable @columnfractions .20 .80
1274 @item @emph{Prototype}: @tab @code{int omp_get_num_teams(void);}
1277 @item @emph{Fortran}:
1278 @multitable @columnfractions .20 .80
1279 @item @emph{Interface}: @tab @code{integer function omp_get_num_teams()}
1282 @item @emph{Reference}:
1283 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.32.
1288 @node omp_get_team_num
1289 @subsection @code{omp_get_team_num} -- Get team number
1291 @item @emph{Description}:
1292 Returns the team number of the calling thread.
1295 @multitable @columnfractions .20 .80
1296 @item @emph{Prototype}: @tab @code{int omp_get_team_num(void);}
1299 @item @emph{Fortran}:
1300 @multitable @columnfractions .20 .80
1301 @item @emph{Interface}: @tab @code{integer function omp_get_team_num()}
1304 @item @emph{Reference}:
1305 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.33.
1310 @node omp_set_num_teams
1311 @subsection @code{omp_set_num_teams} -- Set upper teams limit for teams construct
1313 @item @emph{Description}:
1314 Specifies the upper bound for number of teams created by the teams construct
1315 which does not specify a @code{num_teams} clause. The
1316 argument of @code{omp_set_num_teams} shall be a positive integer.
1319 @multitable @columnfractions .20 .80
1320 @item @emph{Prototype}: @tab @code{void omp_set_num_teams(int num_teams);}
1323 @item @emph{Fortran}:
1324 @multitable @columnfractions .20 .80
1325 @item @emph{Interface}: @tab @code{subroutine omp_set_num_teams(num_teams)}
1326 @item @tab @code{integer, intent(in) :: num_teams}
1329 @item @emph{See also}:
1330 @ref{OMP_NUM_TEAMS}, @ref{omp_get_num_teams}, @ref{omp_get_max_teams}
1332 @item @emph{Reference}:
1333 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.4.3.
1338 @node omp_get_max_teams
1339 @subsection @code{omp_get_max_teams} -- Maximum number of teams of teams region
1341 @item @emph{Description}:
1342 Return the maximum number of teams used for the teams region
1343 that does not use the clause @code{num_teams}.
1346 @multitable @columnfractions .20 .80
1347 @item @emph{Prototype}: @tab @code{int omp_get_max_teams(void);}
1350 @item @emph{Fortran}:
1351 @multitable @columnfractions .20 .80
1352 @item @emph{Interface}: @tab @code{integer function omp_get_max_teams()}
1355 @item @emph{See also}:
1356 @ref{omp_set_num_teams}, @ref{omp_get_num_teams}
1358 @item @emph{Reference}:
1359 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.4.4.
1364 @node omp_set_teams_thread_limit
1365 @subsection @code{omp_set_teams_thread_limit} -- Set upper thread limit for teams construct
1367 @item @emph{Description}:
1368 Specifies the upper bound for number of threads that are available
1369 for each team created by the teams construct which does not specify a
1370 @code{thread_limit} clause. The argument of
1371 @code{omp_set_teams_thread_limit} shall be a positive integer.
1374 @multitable @columnfractions .20 .80
1375 @item @emph{Prototype}: @tab @code{void omp_set_teams_thread_limit(int thread_limit);}
1378 @item @emph{Fortran}:
1379 @multitable @columnfractions .20 .80
1380 @item @emph{Interface}: @tab @code{subroutine omp_set_teams_thread_limit(thread_limit)}
1381 @item @tab @code{integer, intent(in) :: thread_limit}
1384 @item @emph{See also}:
1385 @ref{OMP_TEAMS_THREAD_LIMIT}, @ref{omp_get_teams_thread_limit}, @ref{omp_get_thread_limit}
1387 @item @emph{Reference}:
1388 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.4.5.
1393 @node omp_get_thread_limit
1394 @subsection @code{omp_get_thread_limit} -- Maximum number of threads
1396 @item @emph{Description}:
1397 Return the maximum number of threads of the program.
1400 @multitable @columnfractions .20 .80
1401 @item @emph{Prototype}: @tab @code{int omp_get_thread_limit(void);}
1404 @item @emph{Fortran}:
1405 @multitable @columnfractions .20 .80
1406 @item @emph{Interface}: @tab @code{integer function omp_get_thread_limit()}
1409 @item @emph{See also}:
1410 @ref{omp_get_max_threads}, @ref{OMP_THREAD_LIMIT}
1412 @item @emph{Reference}:
1413 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.14.
1418 @node Tasking Routines
1419 @section Tasking Routines
1421 Routines relating to explicit tasks.
1422 They have C linkage and do not throw exceptions.
1425 * omp_get_max_task_priority:: Maximum task priority value that can be set
1426 * omp_in_explicit_task:: Whether a given task is an explicit task
1427 * omp_in_final:: Whether in final or included task region
1428 @c * omp_is_free_agent:: <fixme>/TR12
1429 @c * omp_ancestor_is_free_agent:: <fixme>/TR12
1434 @node omp_get_max_task_priority
1435 @subsection @code{omp_get_max_task_priority} -- Maximum priority value
1436 that can be set for tasks.
1438 @item @emph{Description}:
1439 This function obtains the maximum allowed priority number for tasks.
1442 @multitable @columnfractions .20 .80
1443 @item @emph{Prototype}: @tab @code{int omp_get_max_task_priority(void);}
1446 @item @emph{Fortran}:
1447 @multitable @columnfractions .20 .80
1448 @item @emph{Interface}: @tab @code{integer function omp_get_max_task_priority()}
1451 @item @emph{Reference}:
1452 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.29.
1457 @node omp_in_explicit_task
1458 @subsection @code{omp_in_explicit_task} -- Whether a given task is an explicit task
1460 @item @emph{Description}:
1461 The function returns the @var{explicit-task-var} ICV; it returns true when the
1462 encountering task was generated by a task-generating construct such as
1463 @code{target}, @code{task} or @code{taskloop}. Otherwise, the encountering task
1464 is in an implicit task region such as generated by the implicit or explicit
1465 @code{parallel} region and @code{omp_in_explicit_task} returns false.
1468 @multitable @columnfractions .20 .80
1469 @item @emph{Prototype}: @tab @code{int omp_in_explicit_task(void);}
1472 @item @emph{Fortran}:
1473 @multitable @columnfractions .20 .80
1474 @item @emph{Interface}: @tab @code{logical function omp_in_explicit_task()}
1477 @item @emph{Reference}:
1478 @uref{https://www.openmp.org, OpenMP specification v5.2}, Section 18.5.2.
1484 @subsection @code{omp_in_final} -- Whether in final or included task region
1486 @item @emph{Description}:
1487 This function returns @code{true} if currently running in a final
1488 or included task region, @code{false} otherwise. Here, @code{true}
1489 and @code{false} represent their language-specific counterparts.
1492 @multitable @columnfractions .20 .80
1493 @item @emph{Prototype}: @tab @code{int omp_in_final(void);}
1496 @item @emph{Fortran}:
1497 @multitable @columnfractions .20 .80
1498 @item @emph{Interface}: @tab @code{logical function omp_in_final()}
1501 @item @emph{Reference}:
1502 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.21.
1507 @node Resource Relinquishing Routines
1508 @section Resource Relinquishing Routines
1510 Routines releasing resources used by the OpenMP runtime.
1511 They have C linkage and do not throw exceptions.
1514 * omp_pause_resource:: Release OpenMP resources on a device
1515 * omp_pause_resource_all:: Release OpenMP resources on all devices
1520 @node omp_pause_resource
1521 @subsection @code{omp_pause_resource} -- Release OpenMP resources on a device
1523 @item @emph{Description}:
1524 Free resources used by the OpenMP program and the runtime library on and for the
1525 device specified by @var{device_num}; on success, zero is returned and non-zero
1528 The value of @var{device_num} must be a conforming device number. The routine
1529 may not be called from within any explicit region and all explicit threads that
1530 do not bind to the implicit parallel region have finalized execution.
1533 @multitable @columnfractions .20 .80
1534 @item @emph{Prototype}: @tab @code{int omp_pause_resource(omp_pause_resource_t kind, int device_num);}
1537 @item @emph{Fortran}:
1538 @multitable @columnfractions .20 .80
1539 @item @emph{Interface}: @tab @code{integer function omp_pause_resource(kind, device_num)}
1540 @item @tab @code{integer (kind=omp_pause_resource_kind) kind}
1541 @item @tab @code{integer device_num}
1544 @item @emph{Reference}:
1545 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.2.43.
1550 @node omp_pause_resource_all
1551 @subsection @code{omp_pause_resource_all} -- Release OpenMP resources on all devices
1553 @item @emph{Description}:
1554 Free resources used by the OpenMP program and the runtime library on all devices,
1555 including the host. On success, zero is returned and non-zero otherwise.
1557 The routine may not be called from within any explicit region and all explicit
1558 threads that do not bind to the implicit parallel region have finalized execution.
1561 @multitable @columnfractions .20 .80
1562 @item @emph{Prototype}: @tab @code{int omp_pause_resource(omp_pause_resource_t kind);}
1565 @item @emph{Fortran}:
1566 @multitable @columnfractions .20 .80
1567 @item @emph{Interface}: @tab @code{integer function omp_pause_resource(kind)}
1568 @item @tab @code{integer (kind=omp_pause_resource_kind) kind}
1571 @item @emph{See also}:
1572 @ref{omp_pause_resource}
1574 @item @emph{Reference}:
1575 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.2.44.
1580 @node Device Information Routines
1581 @section Device Information Routines
1583 Routines related to devices available to an OpenMP program.
1584 They have C linkage and do not throw exceptions.
1587 * omp_get_num_procs:: Number of processors online
1588 @c * omp_get_max_progress_width:: <fixme>/TR11
1589 * omp_set_default_device:: Set the default device for target regions
1590 * omp_get_default_device:: Get the default device for target regions
1591 * omp_get_num_devices:: Number of target devices
1592 * omp_get_device_num:: Get device that current thread is running on
1593 * omp_is_initial_device:: Whether executing on the host device
1594 * omp_get_initial_device:: Device number of host device
1599 @node omp_get_num_procs
1600 @subsection @code{omp_get_num_procs} -- Number of processors online
1602 @item @emph{Description}:
1603 Returns the number of processors online on that device.
1606 @multitable @columnfractions .20 .80
1607 @item @emph{Prototype}: @tab @code{int omp_get_num_procs(void);}
1610 @item @emph{Fortran}:
1611 @multitable @columnfractions .20 .80
1612 @item @emph{Interface}: @tab @code{integer function omp_get_num_procs()}
1615 @item @emph{Reference}:
1616 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.5.
1621 @node omp_set_default_device
1622 @subsection @code{omp_set_default_device} -- Set the default device for target regions
1624 @item @emph{Description}:
1625 Set the default device for target regions without device clause. The argument
1626 shall be a nonnegative device number.
1629 @multitable @columnfractions .20 .80
1630 @item @emph{Prototype}: @tab @code{void omp_set_default_device(int device_num);}
1633 @item @emph{Fortran}:
1634 @multitable @columnfractions .20 .80
1635 @item @emph{Interface}: @tab @code{subroutine omp_set_default_device(device_num)}
1636 @item @tab @code{integer device_num}
1639 @item @emph{See also}:
1640 @ref{OMP_DEFAULT_DEVICE}, @ref{omp_get_default_device}
1642 @item @emph{Reference}:
1643 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.29.
1648 @node omp_get_default_device
1649 @subsection @code{omp_get_default_device} -- Get the default device for target regions
1651 @item @emph{Description}:
1652 Get the default device for target regions without device clause.
1655 @multitable @columnfractions .20 .80
1656 @item @emph{Prototype}: @tab @code{int omp_get_default_device(void);}
1659 @item @emph{Fortran}:
1660 @multitable @columnfractions .20 .80
1661 @item @emph{Interface}: @tab @code{integer function omp_get_default_device()}
1664 @item @emph{See also}:
1665 @ref{OMP_DEFAULT_DEVICE}, @ref{omp_set_default_device}
1667 @item @emph{Reference}:
1668 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.30.
1673 @node omp_get_num_devices
1674 @subsection @code{omp_get_num_devices} -- Number of target devices
1676 @item @emph{Description}:
1677 Returns the number of target devices.
1680 @multitable @columnfractions .20 .80
1681 @item @emph{Prototype}: @tab @code{int omp_get_num_devices(void);}
1684 @item @emph{Fortran}:
1685 @multitable @columnfractions .20 .80
1686 @item @emph{Interface}: @tab @code{integer function omp_get_num_devices()}
1689 @item @emph{Reference}:
1690 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.31.
1695 @node omp_get_device_num
1696 @subsection @code{omp_get_device_num} -- Return device number of current device
1698 @item @emph{Description}:
1699 This function returns a device number that represents the device that the
1700 current thread is executing on. For OpenMP 5.0, this must be equal to the
1701 value returned by the @code{omp_get_initial_device} function when called
1705 @multitable @columnfractions .20 .80
1706 @item @emph{Prototype}: @tab @code{int omp_get_device_num(void);}
1709 @item @emph{Fortran}:
1710 @multitable @columnfractions .20 .80
1711 @item @emph{Interface}: @tab @code{integer function omp_get_device_num()}
1714 @item @emph{See also}:
1715 @ref{omp_get_initial_device}
1717 @item @emph{Reference}:
1718 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.2.37.
1723 @node omp_is_initial_device
1724 @subsection @code{omp_is_initial_device} -- Whether executing on the host device
1726 @item @emph{Description}:
1727 This function returns @code{true} if currently running on the host device,
1728 @code{false} otherwise. Here, @code{true} and @code{false} represent
1729 their language-specific counterparts.
1732 @multitable @columnfractions .20 .80
1733 @item @emph{Prototype}: @tab @code{int omp_is_initial_device(void);}
1736 @item @emph{Fortran}:
1737 @multitable @columnfractions .20 .80
1738 @item @emph{Interface}: @tab @code{logical function omp_is_initial_device()}
1741 @item @emph{Reference}:
1742 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.34.
1747 @node omp_get_initial_device
1748 @subsection @code{omp_get_initial_device} -- Return device number of initial device
1750 @item @emph{Description}:
1751 This function returns a device number that represents the host device.
1752 For OpenMP 5.1, this must be equal to the value returned by the
1753 @code{omp_get_num_devices} function.
1756 @multitable @columnfractions .20 .80
1757 @item @emph{Prototype}: @tab @code{int omp_get_initial_device(void);}
1760 @item @emph{Fortran}:
1761 @multitable @columnfractions .20 .80
1762 @item @emph{Interface}: @tab @code{integer function omp_get_initial_device()}
1765 @item @emph{See also}:
1766 @ref{omp_get_num_devices}
1768 @item @emph{Reference}:
1769 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.35.
1774 @node Device Memory Routines
1775 @section Device Memory Routines
1777 Routines related to memory allocation and managing corresponding
1778 pointers on devices. They have C linkage and do not throw exceptions.
1781 * omp_target_alloc:: Allocate device memory
1782 * omp_target_free:: Free device memory
1783 * omp_target_is_present:: Check whether storage is mapped
1784 * omp_target_is_accessible:: Check whether memory is device accessible
1785 * omp_target_memcpy:: Copy data between devices
1786 * omp_target_memcpy_rect:: Copy a subvolume of data between devices
1787 * omp_target_memcpy_async:: Copy data between devices asynchronously
1788 * omp_target_memcpy_rect_async:: Copy a subvolume of data between devices asynchronously
1789 @c * omp_target_memset:: <fixme>/TR12
1790 @c * omp_target_memset_async:: <fixme>/TR12
1791 * omp_target_associate_ptr:: Associate a device pointer with a host pointer
1792 * omp_target_disassociate_ptr:: Remove device--host pointer association
1793 * omp_get_mapped_ptr:: Return device pointer to a host pointer
1798 @node omp_target_alloc
1799 @subsection @code{omp_target_alloc} -- Allocate device memory
1801 @item @emph{Description}:
1802 This routine allocates @var{size} bytes of memory in the device environment
1803 associated with the device number @var{device_num}. If successful, a device
1804 pointer is returned, otherwise a null pointer.
1806 In GCC, when the device is the host or the device shares memory with the host,
1807 the memory is allocated on the host; in that case, when @var{size} is zero,
1808 either NULL or a unique pointer value that can later be successfully passed to
1809 @code{omp_target_free} is returned. When the allocation is not performed on
1810 the host, a null pointer is returned when @var{size} is zero; in that case,
1811 additionally a diagnostic might be printed to standard error (stderr).
1813 Running this routine in a @code{target} region except on the initial device
1817 @multitable @columnfractions .20 .80
1818 @item @emph{Prototype}: @tab @code{void *omp_target_alloc(size_t size, int device_num)}
1821 @item @emph{Fortran}:
1822 @multitable @columnfractions .20 .80
1823 @item @emph{Interface}: @tab @code{type(c_ptr) function omp_target_alloc(size, device_num) bind(C)}
1824 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_int, c_size_t}
1825 @item @tab @code{integer(c_size_t), value :: size}
1826 @item @tab @code{integer(c_int), value :: device_num}
1829 @item @emph{See also}:
1830 @ref{omp_target_free}, @ref{omp_target_associate_ptr}
1832 @item @emph{Reference}:
1833 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.1
1838 @node omp_target_free
1839 @subsection @code{omp_target_free} -- Free device memory
1841 @item @emph{Description}:
1842 This routine frees memory allocated by the @code{omp_target_alloc} routine.
1843 The @var{device_ptr} argument must be either a null pointer or a device pointer
1844 returned by @code{omp_target_alloc} for the specified @code{device_num}. The
1845 device number @var{device_num} must be a conforming device number.
1847 Running this routine in a @code{target} region except on the initial device
1851 @multitable @columnfractions .20 .80
1852 @item @emph{Prototype}: @tab @code{void omp_target_free(void *device_ptr, int device_num)}
1855 @item @emph{Fortran}:
1856 @multitable @columnfractions .20 .80
1857 @item @emph{Interface}: @tab @code{subroutine omp_target_free(device_ptr, device_num) bind(C)}
1858 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_int}
1859 @item @tab @code{type(c_ptr), value :: device_ptr}
1860 @item @tab @code{integer(c_int), value :: device_num}
1863 @item @emph{See also}:
1864 @ref{omp_target_alloc}, @ref{omp_target_disassociate_ptr}
1866 @item @emph{Reference}:
1867 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.2
1872 @node omp_target_is_present
1873 @subsection @code{omp_target_is_present} -- Check whether storage is mapped
1875 @item @emph{Description}:
1876 This routine tests whether storage, identified by the host pointer @var{ptr}
1877 is mapped to the device specified by @var{device_num}. If so, it returns
1878 a nonzero value and otherwise zero.
1880 In GCC, this includes self mapping such that @code{omp_target_is_present}
1881 returns @emph{true} when @var{device_num} specifies the host or when the host
1882 and the device share memory. If @var{ptr} is a null pointer, @var{true} is
1883 returned and if @var{device_num} is an invalid device number, @var{false} is
1886 If those conditions do not apply, @emph{true} is returned if the association has
1887 been established by an explicit or implicit @code{map} clause, the
1888 @code{declare target} directive or a call to the @code{omp_target_associate_ptr}
1891 Running this routine in a @code{target} region except on the initial device
1895 @multitable @columnfractions .20 .80
1896 @item @emph{Prototype}: @tab @code{int omp_target_is_present(const void *ptr,}
1897 @item @tab @code{ int device_num)}
1900 @item @emph{Fortran}:
1901 @multitable @columnfractions .20 .80
1902 @item @emph{Interface}: @tab @code{integer(c_int) function omp_target_is_present(ptr, &}
1903 @item @tab @code{ device_num) bind(C)}
1904 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_int}
1905 @item @tab @code{type(c_ptr), value :: ptr}
1906 @item @tab @code{integer(c_int), value :: device_num}
1909 @item @emph{See also}:
1910 @ref{omp_target_associate_ptr}
1912 @item @emph{Reference}:
1913 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.3
1918 @node omp_target_is_accessible
1919 @subsection @code{omp_target_is_accessible} -- Check whether memory is device accessible
1921 @item @emph{Description}:
1922 This routine tests whether memory, starting at the address given by @var{ptr}
1923 and extending @var{size} bytes, is accessibly on the device specified by
1924 @var{device_num}. If so, it returns a nonzero value and otherwise zero.
1926 The address given by @var{ptr} is interpreted to be in the address space of
1927 the device and @var{size} must be positive.
1929 Note that GCC's current implementation assumes that @var{ptr} is a valid host
1930 pointer. Therefore, all addresses given by @var{ptr} are assumed to be
1931 accessible on the initial device. And, to err on the safe side, this memory
1932 is only available on a non-host device that can access all host memory
1933 ([uniform] shared memory access).
1935 Running this routine in a @code{target} region except on the initial device
1939 @multitable @columnfractions .20 .80
1940 @item @emph{Prototype}: @tab @code{int omp_target_is_accessible(const void *ptr,}
1941 @item @tab @code{ size_t size,}
1942 @item @tab @code{ int device_num)}
1945 @item @emph{Fortran}:
1946 @multitable @columnfractions .20 .80
1947 @item @emph{Interface}: @tab @code{integer(c_int) function omp_target_is_accessible(ptr, &}
1948 @item @tab @code{ size, device_num) bind(C)}
1949 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_size_t, c_int}
1950 @item @tab @code{type(c_ptr), value :: ptr}
1951 @item @tab @code{integer(c_size_t), value :: size}
1952 @item @tab @code{integer(c_int), value :: device_num}
1955 @item @emph{See also}:
1956 @ref{omp_target_associate_ptr}
1958 @item @emph{Reference}:
1959 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.4
1964 @node omp_target_memcpy
1965 @subsection @code{omp_target_memcpy} -- Copy data between devices
1967 @item @emph{Description}:
1968 This routine copies @var{length} of bytes of data from the device
1969 identified by device number @var{src_device_num} to device @var{dst_device_num}.
1970 The data is copied from the source device from the address provided by
1971 @var{src}, shifted by the offset of @var{src_offset} bytes, to the destination
1972 device's @var{dst} address shifted by @var{dst_offset}. The routine returns
1973 zero on success and non-zero otherwise.
1975 Running this routine in a @code{target} region except on the initial device
1979 @multitable @columnfractions .20 .80
1980 @item @emph{Prototype}: @tab @code{int omp_target_memcpy(void *dst,}
1981 @item @tab @code{ const void *src,}
1982 @item @tab @code{ size_t length,}
1983 @item @tab @code{ size_t dst_offset,}
1984 @item @tab @code{ size_t src_offset,}
1985 @item @tab @code{ int dst_device_num,}
1986 @item @tab @code{ int src_device_num)}
1989 @item @emph{Fortran}:
1990 @multitable @columnfractions .20 .80
1991 @item @emph{Interface}: @tab @code{integer(c_int) function omp_target_memcpy( &}
1992 @item @tab @code{ dst, src, length, dst_offset, src_offset, &}
1993 @item @tab @code{ dst_device_num, src_device_num) bind(C)}
1994 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_size_t, c_int}
1995 @item @tab @code{type(c_ptr), value :: dst, src}
1996 @item @tab @code{integer(c_size_t), value :: length, dst_offset, src_offset}
1997 @item @tab @code{integer(c_int), value :: dst_device_num, src_device_num}
2000 @item @emph{See also}:
2001 @ref{omp_target_memcpy_async}, @ref{omp_target_memcpy_rect}
2003 @item @emph{Reference}:
2004 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.5
2009 @node omp_target_memcpy_async
2010 @subsection @code{omp_target_memcpy_async} -- Copy data between devices asynchronously
2012 @item @emph{Description}:
2013 This routine copies asynchronously @var{length} of bytes of data from the
2014 device identified by device number @var{src_device_num} to device
2015 @var{dst_device_num}. The data is copied from the source device from the
2016 address provided by @var{src}, shifted by the offset of @var{src_offset} bytes,
2017 to the destination device's @var{dst} address shifted by @var{dst_offset}.
2018 Task dependence is expressed by passing an array of depend objects to
2019 @var{depobj_list}, where the number of array elements is passed as
2020 @var{depobj_count}; if the count is zero, the @var{depobj_list} argument is
2021 ignored. The routine returns zero if the copying process has successfully
2022 been started and non-zero otherwise.
2024 Running this routine in a @code{target} region except on the initial device
2028 @multitable @columnfractions .20 .80
2029 @item @emph{Prototype}: @tab @code{int omp_target_memcpy_async(void *dst,}
2030 @item @tab @code{ const void *src,}
2031 @item @tab @code{ size_t length,}
2032 @item @tab @code{ size_t dst_offset,}
2033 @item @tab @code{ size_t src_offset,}
2034 @item @tab @code{ int dst_device_num,}
2035 @item @tab @code{ int src_device_num,}
2036 @item @tab @code{ int depobj_count,}
2037 @item @tab @code{ omp_depend_t *depobj_list)}
2040 @item @emph{Fortran}:
2041 @multitable @columnfractions .20 .80
2042 @item @emph{Interface}: @tab @code{integer(c_int) function omp_target_memcpy_async( &}
2043 @item @tab @code{ dst, src, length, dst_offset, src_offset, &}
2044 @item @tab @code{ dst_device_num, src_device_num, &}
2045 @item @tab @code{ depobj_count, depobj_list) bind(C)}
2046 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_size_t, c_int}
2047 @item @tab @code{type(c_ptr), value :: dst, src}
2048 @item @tab @code{integer(c_size_t), value :: length, dst_offset, src_offset}
2049 @item @tab @code{integer(c_int), value :: dst_device_num, src_device_num, depobj_count}
2050 @item @tab @code{integer(omp_depend_kind), optional :: depobj_list(*)}
2053 @item @emph{See also}:
2054 @ref{omp_target_memcpy}, @ref{omp_target_memcpy_rect_async}
2056 @item @emph{Reference}:
2057 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.7
2062 @node omp_target_memcpy_rect
2063 @subsection @code{omp_target_memcpy_rect} -- Copy a subvolume of data between devices
2065 @item @emph{Description}:
2066 This routine copies a subvolume of data from the device identified by
2067 device number @var{src_device_num} to device @var{dst_device_num}.
2068 The array has @var{num_dims} dimensions and each array element has a size of
2069 @var{element_size} bytes. The @var{volume} array specifies how many elements
2070 per dimension are copied. The full sizes of the destination and source arrays
2071 are given by the @var{dst_dimensions} and @var{src_dimensions} arguments,
2072 respectively. The offset per dimension to the first element to be copied is
2073 given by the @var{dst_offset} and @var{src_offset} arguments. The routine
2074 returns zero on success and non-zero otherwise.
2076 The OpenMP specification only requires that @var{num_dims} up to three is
2077 supported. In order to find implementation-specific maximally supported number
2078 of dimensions, the routine returns this value when invoked with a null pointer
2079 to both the @var{dst} and @var{src} arguments. As GCC supports arbitrary
2080 dimensions, it returns @code{INT_MAX}.
2082 The device-number arguments must be conforming device numbers, the @var{src} and
2083 @var{dst} must be either both null pointers or all of the following must be
2084 fulfilled: @var{element_size} and @var{num_dims} must be positive and the
2085 @var{volume}, offset and dimension arrays must have at least @var{num_dims}
2088 Running this routine in a @code{target} region is not supported except on
2092 @multitable @columnfractions .20 .80
2093 @item @emph{Prototype}: @tab @code{int omp_target_memcpy_rect(void *dst,}
2094 @item @tab @code{ const void *src,}
2095 @item @tab @code{ size_t element_size,}
2096 @item @tab @code{ int num_dims,}
2097 @item @tab @code{ const size_t *volume,}
2098 @item @tab @code{ const size_t *dst_offset,}
2099 @item @tab @code{ const size_t *src_offset,}
2100 @item @tab @code{ const size_t *dst_dimensions,}
2101 @item @tab @code{ const size_t *src_dimensions,}
2102 @item @tab @code{ int dst_device_num,}
2103 @item @tab @code{ int src_device_num)}
2106 @item @emph{Fortran}:
2107 @multitable @columnfractions .20 .80
2108 @item @emph{Interface}: @tab @code{integer(c_int) function omp_target_memcpy_rect( &}
2109 @item @tab @code{ dst, src, element_size, num_dims, volume, &}
2110 @item @tab @code{ dst_offset, src_offset, dst_dimensions, &}
2111 @item @tab @code{ src_dimensions, dst_device_num, src_device_num) bind(C)}
2112 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_size_t, c_int}
2113 @item @tab @code{type(c_ptr), value :: dst, src}
2114 @item @tab @code{integer(c_size_t), value :: element_size, dst_offset, src_offset}
2115 @item @tab @code{integer(c_size_t), value :: volume, dst_dimensions, src_dimensions}
2116 @item @tab @code{integer(c_int), value :: num_dims, dst_device_num, src_device_num}
2119 @item @emph{See also}:
2120 @ref{omp_target_memcpy_rect_async}, @ref{omp_target_memcpy}
2122 @item @emph{Reference}:
2123 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.6
2128 @node omp_target_memcpy_rect_async
2129 @subsection @code{omp_target_memcpy_rect_async} -- Copy a subvolume of data between devices asynchronously
2131 @item @emph{Description}:
2132 This routine copies asynchronously a subvolume of data from the device
2133 identified by device number @var{src_device_num} to device @var{dst_device_num}.
2134 The array has @var{num_dims} dimensions and each array element has a size of
2135 @var{element_size} bytes. The @var{volume} array specifies how many elements
2136 per dimension are copied. The full sizes of the destination and source arrays
2137 are given by the @var{dst_dimensions} and @var{src_dimensions} arguments,
2138 respectively. The offset per dimension to the first element to be copied is
2139 given by the @var{dst_offset} and @var{src_offset} arguments. Task dependence
2140 is expressed by passing an array of depend objects to @var{depobj_list}, where
2141 the number of array elements is passed as @var{depobj_count}; if the count is
2142 zero, the @var{depobj_list} argument is ignored. The routine
2143 returns zero on success and non-zero otherwise.
2145 The OpenMP specification only requires that @var{num_dims} up to three is
2146 supported. In order to find implementation-specific maximally supported number
2147 of dimensions, the routine returns this value when invoked with a null pointer
2148 to both the @var{dst} and @var{src} arguments. As GCC supports arbitrary
2149 dimensions, it returns @code{INT_MAX}.
2151 The device-number arguments must be conforming device numbers, the @var{src} and
2152 @var{dst} must be either both null pointers or all of the following must be
2153 fulfilled: @var{element_size} and @var{num_dims} must be positive and the
2154 @var{volume}, offset and dimension arrays must have at least @var{num_dims}
2157 Running this routine in a @code{target} region is not supported except on
2163 @multitable @columnfractions .20 .80
2164 @item @emph{Prototype}: @tab @code{int omp_target_memcpy_rect_async(void *dst,}
2165 @item @tab @code{ const void *src,}
2166 @item @tab @code{ size_t element_size,}
2167 @item @tab @code{ int num_dims,}
2168 @item @tab @code{ const size_t *volume,}
2169 @item @tab @code{ const size_t *dst_offset,}
2170 @item @tab @code{ const size_t *src_offset,}
2171 @item @tab @code{ const size_t *dst_dimensions,}
2172 @item @tab @code{ const size_t *src_dimensions,}
2173 @item @tab @code{ int dst_device_num,}
2174 @item @tab @code{ int src_device_num,}
2175 @item @tab @code{ int depobj_count,}
2176 @item @tab @code{ omp_depend_t *depobj_list)}
2179 @item @emph{Fortran}:
2180 @multitable @columnfractions .20 .80
2181 @item @emph{Interface}: @tab @code{integer(c_int) function omp_target_memcpy_rect_async( &}
2182 @item @tab @code{ dst, src, element_size, num_dims, volume, &}
2183 @item @tab @code{ dst_offset, src_offset, dst_dimensions, &}
2184 @item @tab @code{ src_dimensions, dst_device_num, src_device_num, &}
2185 @item @tab @code{ depobj_count, depobj_list) bind(C)}
2186 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_size_t, c_int}
2187 @item @tab @code{type(c_ptr), value :: dst, src}
2188 @item @tab @code{integer(c_size_t), value :: element_size, dst_offset, src_offset}
2189 @item @tab @code{integer(c_size_t), value :: volume, dst_dimensions, src_dimensions}
2190 @item @tab @code{integer(c_int), value :: num_dims, dst_device_num, src_device_num}
2191 @item @tab @code{integer(c_int), value :: depobj_count}
2192 @item @tab @code{integer(omp_depend_kind), optional :: depobj_list(*)}
2195 @item @emph{See also}:
2196 @ref{omp_target_memcpy_rect}, @ref{omp_target_memcpy_async}
2198 @item @emph{Reference}:
2199 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.8
2204 @node omp_target_associate_ptr
2205 @subsection @code{omp_target_associate_ptr} -- Associate a device pointer with a host pointer
2207 @item @emph{Description}:
2208 This routine associates storage on the host with storage on a device identified
2209 by @var{device_num}. The device pointer is usually obtained by calling
2210 @code{omp_target_alloc} or by other means (but not by using the @code{map}
2211 clauses or the @code{declare target} directive). The host pointer should point
2212 to memory that has a storage size of at least @var{size}.
2214 The @var{device_offset} parameter specifies the offset into @var{device_ptr}
2215 that is used as the base address for the device side of the mapping; the
2216 storage size should be at least @var{device_offset} plus @var{size}.
2218 After the association, the host pointer can be used in a @code{map} clause and
2219 in the @code{to} and @code{from} clauses of the @code{target update} directive
2220 to transfer data between the associated pointers. The reference count of such
2221 associated storage is infinite. The association can be removed by calling
2222 @code{omp_target_disassociate_ptr} which should be done before the lifetime
2223 of either either storage ends.
2225 The routine returns nonzero (@code{EINVAL}) when the @var{device_num} invalid,
2226 for when the initial device or the associated device shares memory with the
2227 host. @code{omp_target_associate_ptr} returns zero if @var{host_ptr} points
2228 into already associated storage that is fully inside of a previously associated
2229 memory. Otherwise, if the association was successful zero is returned; if none
2230 of the cases above apply, nonzero (@code{EINVAL}) is returned.
2232 The @code{omp_target_is_present} routine can be used to test whether
2233 associated storage for a device pointer exists.
2235 Running this routine in a @code{target} region except on the initial device
2239 @multitable @columnfractions .20 .80
2240 @item @emph{Prototype}: @tab @code{int omp_target_associate_ptr(const void *host_ptr,}
2241 @item @tab @code{ const void *device_ptr,}
2242 @item @tab @code{ size_t size,}
2243 @item @tab @code{ size_t device_offset,}
2244 @item @tab @code{ int device_num)}
2247 @item @emph{Fortran}:
2248 @multitable @columnfractions .20 .80
2249 @item @emph{Interface}: @tab @code{integer(c_int) function omp_target_associate_ptr(host_ptr, &}
2250 @item @tab @code{ device_ptr, size, device_offset, device_num) bind(C)}
2251 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_int, c_size_t}
2252 @item @tab @code{type(c_ptr), value :: host_ptr, device_ptr}
2253 @item @tab @code{integer(c_size_t), value :: size, device_offset}
2254 @item @tab @code{integer(c_int), value :: device_num}
2257 @item @emph{See also}:
2258 @ref{omp_target_disassociate_ptr}, @ref{omp_target_is_present},
2259 @ref{omp_target_alloc}
2261 @item @emph{Reference}:
2262 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.9
2267 @node omp_target_disassociate_ptr
2268 @subsection @code{omp_target_disassociate_ptr} -- Remove device--host pointer association
2270 @item @emph{Description}:
2271 This routine removes the storage association established by calling
2272 @code{omp_target_associate_ptr} and sets the reference count to zero,
2273 even if @code{omp_target_associate_ptr} was invoked multiple times for
2274 for host pointer @code{ptr}. If applicable, the device memory needs
2275 to be freed by the user.
2277 If an associated device storage location for the @var{device_num} was
2278 found and has infinite reference count, the association is removed and
2279 zero is returned. In all other cases, nonzero (@code{EINVAL}) is returned
2280 and no other action is taken.
2282 Note that passing a host pointer where the association to the device pointer
2283 was established with the @code{declare target} directive yields undefined
2286 Running this routine in a @code{target} region except on the initial device
2290 @multitable @columnfractions .20 .80
2291 @item @emph{Prototype}: @tab @code{int omp_target_disassociate_ptr(const void *ptr,}
2292 @item @tab @code{ int device_num)}
2295 @item @emph{Fortran}:
2296 @multitable @columnfractions .20 .80
2297 @item @emph{Interface}: @tab @code{integer(c_int) function omp_target_disassociate_ptr(ptr, &}
2298 @item @tab @code{ device_num) bind(C)}
2299 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_int}
2300 @item @tab @code{type(c_ptr), value :: ptr}
2301 @item @tab @code{integer(c_int), value :: device_num}
2304 @item @emph{See also}:
2305 @ref{omp_target_associate_ptr}
2307 @item @emph{Reference}:
2308 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.10
2313 @node omp_get_mapped_ptr
2314 @subsection @code{omp_get_mapped_ptr} -- Return device pointer to a host pointer
2316 @item @emph{Description}:
2317 If the device number is refers to the initial device or to a device with
2318 memory accessible from the host (shared memory), the @code{omp_get_mapped_ptr}
2319 routines returns the value of the passed @var{ptr}. Otherwise, if associated
2320 storage to the passed host pointer @var{ptr} exists on device associated with
2321 @var{device_num}, it returns that pointer. In all other cases and in cases of
2322 an error, a null pointer is returned.
2324 The association of storage location is established either via an explicit or
2325 implicit @code{map} clause, the @code{declare target} directive or the
2326 @code{omp_target_associate_ptr} routine.
2328 Running this routine in a @code{target} region except on the initial device
2332 @multitable @columnfractions .20 .80
2333 @item @emph{Prototype}: @tab @code{void *omp_get_mapped_ptr(const void *ptr, int device_num);}
2336 @item @emph{Fortran}:
2337 @multitable @columnfractions .20 .80
2338 @item @emph{Interface}: @tab @code{type(c_ptr) function omp_get_mapped_ptr(ptr, device_num) bind(C)}
2339 @item @tab @code{use, intrinsic :: iso_c_binding, only: c_ptr, c_int}
2340 @item @tab @code{type(c_ptr), value :: ptr}
2341 @item @tab @code{integer(c_int), value :: device_num}
2344 @item @emph{See also}:
2345 @ref{omp_target_associate_ptr}
2347 @item @emph{Reference}:
2348 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.8.11
2354 @section Lock Routines
2356 Initialize, set, test, unset and destroy simple and nested locks.
2357 The routines have C linkage and do not throw exceptions.
2360 * omp_init_lock:: Initialize simple lock
2361 * omp_init_nest_lock:: Initialize nested lock
2362 @c * omp_init_lock_with_hint:: <fixme>
2363 @c * omp_init_nest_lock_with_hint:: <fixme>
2364 * omp_destroy_lock:: Destroy simple lock
2365 * omp_destroy_nest_lock:: Destroy nested lock
2366 * omp_set_lock:: Wait for and set simple lock
2367 * omp_set_nest_lock:: Wait for and set simple lock
2368 * omp_unset_lock:: Unset simple lock
2369 * omp_unset_nest_lock:: Unset nested lock
2370 * omp_test_lock:: Test and set simple lock if available
2371 * omp_test_nest_lock:: Test and set nested lock if available
2377 @subsection @code{omp_init_lock} -- Initialize simple lock
2379 @item @emph{Description}:
2380 Initialize a simple lock. After initialization, the lock is in
2384 @multitable @columnfractions .20 .80
2385 @item @emph{Prototype}: @tab @code{void omp_init_lock(omp_lock_t *lock);}
2388 @item @emph{Fortran}:
2389 @multitable @columnfractions .20 .80
2390 @item @emph{Interface}: @tab @code{subroutine omp_init_lock(svar)}
2391 @item @tab @code{integer(omp_lock_kind), intent(out) :: svar}
2394 @item @emph{See also}:
2395 @ref{omp_destroy_lock}
2397 @item @emph{Reference}:
2398 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.1.
2403 @node omp_init_nest_lock
2404 @subsection @code{omp_init_nest_lock} -- Initialize nested lock
2406 @item @emph{Description}:
2407 Initialize a nested lock. After initialization, the lock is in
2408 an unlocked state and the nesting count is set to zero.
2411 @multitable @columnfractions .20 .80
2412 @item @emph{Prototype}: @tab @code{void omp_init_nest_lock(omp_nest_lock_t *lock);}
2415 @item @emph{Fortran}:
2416 @multitable @columnfractions .20 .80
2417 @item @emph{Interface}: @tab @code{subroutine omp_init_nest_lock(nvar)}
2418 @item @tab @code{integer(omp_nest_lock_kind), intent(out) :: nvar}
2421 @item @emph{See also}:
2422 @ref{omp_destroy_nest_lock}
2424 @item @emph{Reference}:
2425 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.1.
2430 @node omp_destroy_lock
2431 @subsection @code{omp_destroy_lock} -- Destroy simple lock
2433 @item @emph{Description}:
2434 Destroy a simple lock. In order to be destroyed, a simple lock must be
2435 in the unlocked state.
2438 @multitable @columnfractions .20 .80
2439 @item @emph{Prototype}: @tab @code{void omp_destroy_lock(omp_lock_t *lock);}
2442 @item @emph{Fortran}:
2443 @multitable @columnfractions .20 .80
2444 @item @emph{Interface}: @tab @code{subroutine omp_destroy_lock(svar)}
2445 @item @tab @code{integer(omp_lock_kind), intent(inout) :: svar}
2448 @item @emph{See also}:
2451 @item @emph{Reference}:
2452 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.3.
2457 @node omp_destroy_nest_lock
2458 @subsection @code{omp_destroy_nest_lock} -- Destroy nested lock
2460 @item @emph{Description}:
2461 Destroy a nested lock. In order to be destroyed, a nested lock must be
2462 in the unlocked state and its nesting count must equal zero.
2465 @multitable @columnfractions .20 .80
2466 @item @emph{Prototype}: @tab @code{void omp_destroy_nest_lock(omp_nest_lock_t *);}
2469 @item @emph{Fortran}:
2470 @multitable @columnfractions .20 .80
2471 @item @emph{Interface}: @tab @code{subroutine omp_destroy_nest_lock(nvar)}
2472 @item @tab @code{integer(omp_nest_lock_kind), intent(inout) :: nvar}
2475 @item @emph{See also}:
2478 @item @emph{Reference}:
2479 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.3.
2485 @subsection @code{omp_set_lock} -- Wait for and set simple lock
2487 @item @emph{Description}:
2488 Before setting a simple lock, the lock variable must be initialized by
2489 @code{omp_init_lock}. The calling thread is blocked until the lock
2490 is available. If the lock is already held by the current thread,
2494 @multitable @columnfractions .20 .80
2495 @item @emph{Prototype}: @tab @code{void omp_set_lock(omp_lock_t *lock);}
2498 @item @emph{Fortran}:
2499 @multitable @columnfractions .20 .80
2500 @item @emph{Interface}: @tab @code{subroutine omp_set_lock(svar)}
2501 @item @tab @code{integer(omp_lock_kind), intent(inout) :: svar}
2504 @item @emph{See also}:
2505 @ref{omp_init_lock}, @ref{omp_test_lock}, @ref{omp_unset_lock}
2507 @item @emph{Reference}:
2508 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.4.
2513 @node omp_set_nest_lock
2514 @subsection @code{omp_set_nest_lock} -- Wait for and set nested lock
2516 @item @emph{Description}:
2517 Before setting a nested lock, the lock variable must be initialized by
2518 @code{omp_init_nest_lock}. The calling thread is blocked until the lock
2519 is available. If the lock is already held by the current thread, the
2520 nesting count for the lock is incremented.
2523 @multitable @columnfractions .20 .80
2524 @item @emph{Prototype}: @tab @code{void omp_set_nest_lock(omp_nest_lock_t *lock);}
2527 @item @emph{Fortran}:
2528 @multitable @columnfractions .20 .80
2529 @item @emph{Interface}: @tab @code{subroutine omp_set_nest_lock(nvar)}
2530 @item @tab @code{integer(omp_nest_lock_kind), intent(inout) :: nvar}
2533 @item @emph{See also}:
2534 @ref{omp_init_nest_lock}, @ref{omp_unset_nest_lock}
2536 @item @emph{Reference}:
2537 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.4.
2542 @node omp_unset_lock
2543 @subsection @code{omp_unset_lock} -- Unset simple lock
2545 @item @emph{Description}:
2546 A simple lock about to be unset must have been locked by @code{omp_set_lock}
2547 or @code{omp_test_lock} before. In addition, the lock must be held by the
2548 thread calling @code{omp_unset_lock}. Then, the lock becomes unlocked. If one
2549 or more threads attempted to set the lock before, one of them is chosen to,
2550 again, set the lock to itself.
2553 @multitable @columnfractions .20 .80
2554 @item @emph{Prototype}: @tab @code{void omp_unset_lock(omp_lock_t *lock);}
2557 @item @emph{Fortran}:
2558 @multitable @columnfractions .20 .80
2559 @item @emph{Interface}: @tab @code{subroutine omp_unset_lock(svar)}
2560 @item @tab @code{integer(omp_lock_kind), intent(inout) :: svar}
2563 @item @emph{See also}:
2564 @ref{omp_set_lock}, @ref{omp_test_lock}
2566 @item @emph{Reference}:
2567 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.5.
2572 @node omp_unset_nest_lock
2573 @subsection @code{omp_unset_nest_lock} -- Unset nested lock
2575 @item @emph{Description}:
2576 A nested lock about to be unset must have been locked by @code{omp_set_nested_lock}
2577 or @code{omp_test_nested_lock} before. In addition, the lock must be held by the
2578 thread calling @code{omp_unset_nested_lock}. If the nesting count drops to zero, the
2579 lock becomes unlocked. If one ore more threads attempted to set the lock before,
2580 one of them is chosen to, again, set the lock to itself.
2583 @multitable @columnfractions .20 .80
2584 @item @emph{Prototype}: @tab @code{void omp_unset_nest_lock(omp_nest_lock_t *lock);}
2587 @item @emph{Fortran}:
2588 @multitable @columnfractions .20 .80
2589 @item @emph{Interface}: @tab @code{subroutine omp_unset_nest_lock(nvar)}
2590 @item @tab @code{integer(omp_nest_lock_kind), intent(inout) :: nvar}
2593 @item @emph{See also}:
2594 @ref{omp_set_nest_lock}
2596 @item @emph{Reference}:
2597 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.5.
2603 @subsection @code{omp_test_lock} -- Test and set simple lock if available
2605 @item @emph{Description}:
2606 Before setting a simple lock, the lock variable must be initialized by
2607 @code{omp_init_lock}. Contrary to @code{omp_set_lock}, @code{omp_test_lock}
2608 does not block if the lock is not available. This function returns
2609 @code{true} upon success, @code{false} otherwise. Here, @code{true} and
2610 @code{false} represent their language-specific counterparts.
2613 @multitable @columnfractions .20 .80
2614 @item @emph{Prototype}: @tab @code{int omp_test_lock(omp_lock_t *lock);}
2617 @item @emph{Fortran}:
2618 @multitable @columnfractions .20 .80
2619 @item @emph{Interface}: @tab @code{logical function omp_test_lock(svar)}
2620 @item @tab @code{integer(omp_lock_kind), intent(inout) :: svar}
2623 @item @emph{See also}:
2624 @ref{omp_init_lock}, @ref{omp_set_lock}, @ref{omp_set_lock}
2626 @item @emph{Reference}:
2627 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.6.
2632 @node omp_test_nest_lock
2633 @subsection @code{omp_test_nest_lock} -- Test and set nested lock if available
2635 @item @emph{Description}:
2636 Before setting a nested lock, the lock variable must be initialized by
2637 @code{omp_init_nest_lock}. Contrary to @code{omp_set_nest_lock},
2638 @code{omp_test_nest_lock} does not block if the lock is not available.
2639 If the lock is already held by the current thread, the new nesting count
2640 is returned. Otherwise, the return value equals zero.
2643 @multitable @columnfractions .20 .80
2644 @item @emph{Prototype}: @tab @code{int omp_test_nest_lock(omp_nest_lock_t *lock);}
2647 @item @emph{Fortran}:
2648 @multitable @columnfractions .20 .80
2649 @item @emph{Interface}: @tab @code{logical function omp_test_nest_lock(nvar)}
2650 @item @tab @code{integer(omp_nest_lock_kind), intent(inout) :: nvar}
2654 @item @emph{See also}:
2655 @ref{omp_init_lock}, @ref{omp_set_lock}, @ref{omp_set_lock}
2657 @item @emph{Reference}:
2658 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.6.
2663 @node Timing Routines
2664 @section Timing Routines
2666 Portable, thread-based, wall clock timer.
2667 The routines have C linkage and do not throw exceptions.
2670 * omp_get_wtick:: Get timer precision.
2671 * omp_get_wtime:: Elapsed wall clock time.
2677 @subsection @code{omp_get_wtick} -- Get timer precision
2679 @item @emph{Description}:
2680 Gets the timer precision, i.e., the number of seconds between two
2681 successive clock ticks.
2684 @multitable @columnfractions .20 .80
2685 @item @emph{Prototype}: @tab @code{double omp_get_wtick(void);}
2688 @item @emph{Fortran}:
2689 @multitable @columnfractions .20 .80
2690 @item @emph{Interface}: @tab @code{double precision function omp_get_wtick()}
2693 @item @emph{See also}:
2696 @item @emph{Reference}:
2697 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.4.2.
2703 @subsection @code{omp_get_wtime} -- Elapsed wall clock time
2705 @item @emph{Description}:
2706 Elapsed wall clock time in seconds. The time is measured per thread, no
2707 guarantee can be made that two distinct threads measure the same time.
2708 Time is measured from some "time in the past", which is an arbitrary time
2709 guaranteed not to change during the execution of the program.
2712 @multitable @columnfractions .20 .80
2713 @item @emph{Prototype}: @tab @code{double omp_get_wtime(void);}
2716 @item @emph{Fortran}:
2717 @multitable @columnfractions .20 .80
2718 @item @emph{Interface}: @tab @code{double precision function omp_get_wtime()}
2721 @item @emph{See also}:
2724 @item @emph{Reference}:
2725 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.4.1.
2731 @section Event Routine
2733 Support for event objects.
2734 The routine has C linkage and do not throw exceptions.
2737 * omp_fulfill_event:: Fulfill and destroy an OpenMP event.
2742 @node omp_fulfill_event
2743 @subsection @code{omp_fulfill_event} -- Fulfill and destroy an OpenMP event
2745 @item @emph{Description}:
2746 Fulfill the event associated with the event handle argument. Currently, it
2747 is only used to fulfill events generated by detach clauses on task
2748 constructs - the effect of fulfilling the event is to allow the task to
2751 The result of calling @code{omp_fulfill_event} with an event handle other
2752 than that generated by a detach clause is undefined. Calling it with an
2753 event handle that has already been fulfilled is also undefined.
2756 @multitable @columnfractions .20 .80
2757 @item @emph{Prototype}: @tab @code{void omp_fulfill_event(omp_event_handle_t event);}
2760 @item @emph{Fortran}:
2761 @multitable @columnfractions .20 .80
2762 @item @emph{Interface}: @tab @code{subroutine omp_fulfill_event(event)}
2763 @item @tab @code{integer (kind=omp_event_handle_kind) :: event}
2766 @item @emph{Reference}:
2767 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.5.1.
2772 @c @node Interoperability Routines
2773 @c @section Interoperability Routines
2775 @c Routines to obtain properties from an @code{omp_interop_t} object.
2776 @c They have C linkage and do not throw exceptions.
2779 @c * omp_get_num_interop_properties:: <fixme>
2780 @c * omp_get_interop_int:: <fixme>
2781 @c * omp_get_interop_ptr:: <fixme>
2782 @c * omp_get_interop_str:: <fixme>
2783 @c * omp_get_interop_name:: <fixme>
2784 @c * omp_get_interop_type_desc:: <fixme>
2785 @c * omp_get_interop_rc_desc:: <fixme>
2788 @node Memory Management Routines
2789 @section Memory Management Routines
2791 Routines to manage and allocate memory on the current device.
2792 They have C linkage and do not throw exceptions.
2795 * omp_init_allocator:: Create an allocator
2796 * omp_destroy_allocator:: Destroy an allocator
2797 * omp_set_default_allocator:: Set the default allocator
2798 * omp_get_default_allocator:: Get the default allocator
2799 * omp_alloc:: Memory allocation with an allocator
2800 * omp_aligned_alloc:: Memory allocation with an allocator and alignment
2801 * omp_free:: Freeing memory allocated with OpenMP routines
2802 * omp_calloc:: Allocate nullified memory with an allocator
2803 * omp_aligned_calloc:: Allocate nullified aligned memory with an allocator
2804 * omp_realloc:: Reallocate memory allocated with OpenMP routines
2805 @c * omp_get_memspace_num_resources:: <fixme>/TR11
2806 @c * omp_get_submemspace:: <fixme>/TR11
2811 @node omp_init_allocator
2812 @subsection @code{omp_init_allocator} -- Create an allocator
2814 @item @emph{Description}:
2815 Create an allocator that uses the specified memory space and has the specified
2816 traits; if an allocator that fulfills the requirements cannot be created,
2817 @code{omp_null_allocator} is returned.
2819 The predefined memory spaces and available traits can be found at
2820 @ref{OMP_ALLOCATOR}, where the trait names have to be be prefixed by
2821 @code{omp_atk_} (e.g. @code{omp_atk_pinned}) and the named trait values by
2822 @code{omp_atv_} (e.g. @code{omp_atv_true}); additionally, @code{omp_atv_default}
2823 may be used as trait value to specify that the default value should be used.
2826 @multitable @columnfractions .20 .80
2827 @item @emph{Prototype}: @tab @code{omp_allocator_handle_t omp_init_allocator(}
2828 @item @tab @code{ omp_memspace_handle_t memspace,}
2829 @item @tab @code{ int ntraits,}
2830 @item @tab @code{ const omp_alloctrait_t traits[]);}
2833 @item @emph{Fortran}:
2834 @multitable @columnfractions .20 .80
2835 @item @emph{Interface}: @tab @code{function omp_init_allocator(memspace, ntraits, traits)}
2836 @item @tab @code{integer (omp_allocator_handle_kind) :: omp_init_allocator}
2837 @item @tab @code{integer (omp_memspace_handle_kind), intent(in) :: memspace}
2838 @item @tab @code{integer, intent(in) :: ntraits}
2839 @item @tab @code{type (omp_alloctrait), intent(in) :: traits(*)}
2842 @item @emph{See also}:
2843 @ref{OMP_ALLOCATOR}, @ref{Memory allocation}, @ref{omp_destroy_allocator}
2845 @item @emph{Reference}:
2846 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.7.2
2851 @node omp_destroy_allocator
2852 @subsection @code{omp_destroy_allocator} -- Destroy an allocator
2854 @item @emph{Description}:
2855 Releases all resources used by a memory allocator, which must not represent
2856 a predefined memory allocator. Accessing memory after its allocator has been
2857 destroyed has unspecified behavior. Passing @code{omp_null_allocator} to the
2858 routine is permitted but has no effect.
2862 @multitable @columnfractions .20 .80
2863 @item @emph{Prototype}: @tab @code{void omp_destroy_allocator (omp_allocator_handle_t allocator);}
2866 @item @emph{Fortran}:
2867 @multitable @columnfractions .20 .80
2868 @item @emph{Interface}: @tab @code{subroutine omp_destroy_allocator(allocator)}
2869 @item @tab @code{integer (omp_allocator_handle_kind), intent(in) :: allocator}
2872 @item @emph{See also}:
2873 @ref{omp_init_allocator}
2875 @item @emph{Reference}:
2876 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.7.3
2881 @node omp_set_default_allocator
2882 @subsection @code{omp_set_default_allocator} -- Set the default allocator
2884 @item @emph{Description}:
2885 Sets the default allocator that is used when no allocator has been specified
2886 in the @code{allocate} or @code{allocator} clause or if an OpenMP memory
2887 routine is invoked with the @code{omp_null_allocator} allocator.
2890 @multitable @columnfractions .20 .80
2891 @item @emph{Prototype}: @tab @code{void omp_set_default_allocator(omp_allocator_handle_t allocator);}
2894 @item @emph{Fortran}:
2895 @multitable @columnfractions .20 .80
2896 @item @emph{Interface}: @tab @code{subroutine omp_set_default_allocator(allocator)}
2897 @item @tab @code{integer (omp_allocator_handle_kind), intent(in) :: allocator}
2900 @item @emph{See also}:
2901 @ref{omp_get_default_allocator}, @ref{omp_init_allocator}, @ref{OMP_ALLOCATOR},
2902 @ref{Memory allocation}
2904 @item @emph{Reference}:
2905 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.7.4
2910 @node omp_get_default_allocator
2911 @subsection @code{omp_get_default_allocator} -- Get the default allocator
2913 @item @emph{Description}:
2914 The routine returns the default allocator that is used when no allocator has
2915 been specified in the @code{allocate} or @code{allocator} clause or if an
2916 OpenMP memory routine is invoked with the @code{omp_null_allocator} allocator.
2919 @multitable @columnfractions .20 .80
2920 @item @emph{Prototype}: @tab @code{omp_allocator_handle_t omp_get_default_allocator();}
2923 @item @emph{Fortran}:
2924 @multitable @columnfractions .20 .80
2925 @item @emph{Interface}: @tab @code{function omp_get_default_allocator()}
2926 @item @tab @code{integer (omp_allocator_handle_kind) :: omp_get_default_allocator}
2929 @item @emph{See also}:
2930 @ref{omp_set_default_allocator}, @ref{OMP_ALLOCATOR}
2932 @item @emph{Reference}:
2933 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.7.5
2939 @subsection @code{omp_alloc} -- Memory allocation with an allocator
2941 @item @emph{Description}:
2942 Allocate memory with the specified allocator, which can either be a predefined
2943 allocator, an allocator handle or @code{omp_null_allocator}. If the allocators
2944 is @code{omp_null_allocator}, the allocator specified by the
2945 @var{def-allocator-var} ICV is used. @var{size} must be a nonnegative number
2946 denoting the number of bytes to be allocated; if @var{size} is zero,
2947 @code{omp_alloc} will return a null pointer. If successful, a pointer to the
2948 allocated memory is returned, otherwise the @code{fallback} trait of the
2949 allocator determines the behavior. The content of the allocated memory is
2952 In @code{target} regions, either the @code{dynamic_allocators} clause must
2953 appear on a @code{requires} directive in the same compilation unit -- or the
2954 @var{allocator} argument may only be a constant expression with the value of
2955 one of the predefined allocators and may not be @code{omp_null_allocator}.
2957 Memory allocated by @code{omp_alloc} must be freed using @code{omp_free}.
2960 @multitable @columnfractions .20 .80
2961 @item @emph{Prototype}: @tab @code{void* omp_alloc(size_t size,}
2962 @item @tab @code{ omp_allocator_handle_t allocator)}
2966 @multitable @columnfractions .20 .80
2967 @item @emph{Prototype}: @tab @code{void* omp_alloc(size_t size,}
2968 @item @tab @code{ omp_allocator_handle_t allocator=omp_null_allocator)}
2971 @item @emph{Fortran}:
2972 @multitable @columnfractions .20 .80
2973 @item @emph{Interface}: @tab @code{type(c_ptr) function omp_alloc(size, allocator) bind(C)}
2974 @item @tab @code{use, intrinsic :: iso_c_binding, only : c_ptr, c_size_t}
2975 @item @tab @code{integer (c_size_t), value :: size}
2976 @item @tab @code{integer (omp_allocator_handle_kind), value :: allocator}
2979 @item @emph{See also}:
2980 @ref{OMP_ALLOCATOR}, @ref{Memory allocation}, @ref{omp_set_default_allocator},
2981 @ref{omp_free}, @ref{omp_init_allocator}
2983 @item @emph{Reference}:
2984 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.7.6
2989 @node omp_aligned_alloc
2990 @subsection @code{omp_aligned_alloc} -- Memory allocation with an allocator and alignment
2992 @item @emph{Description}:
2993 Allocate memory with the specified allocator, which can either be a predefined
2994 allocator, an allocator handle or @code{omp_null_allocator}. If the allocators
2995 is @code{omp_null_allocator}, the allocator specified by the
2996 @var{def-allocator-var} ICV is used. @var{alignment} must be a positive power
2997 of two and @var{size} must be a nonnegative number that is a multiple of the
2998 alignment and denotes the number of bytes to be allocated; if @var{size} is
2999 zero, @code{omp_aligned_alloc} will return a null pointer. The alignment will
3000 be at least the maximal value required by @code{alignment} trait of the
3001 allocator and the value of the passed @var{alignment} argument. If successful,
3002 a pointer to the allocated memory is returned, otherwise the @code{fallback}
3003 trait of the allocator determines the behavior. The content of the allocated
3004 memory is unspecified.
3006 In @code{target} regions, either the @code{dynamic_allocators} clause must
3007 appear on a @code{requires} directive in the same compilation unit -- or the
3008 @var{allocator} argument may only be a constant expression with the value of
3009 one of the predefined allocators and may not be @code{omp_null_allocator}.
3011 Memory allocated by @code{omp_aligned_alloc} must be freed using
3015 @multitable @columnfractions .20 .80
3016 @item @emph{Prototype}: @tab @code{void* omp_aligned_alloc(size_t alignment,}
3017 @item @tab @code{ size_t size,}
3018 @item @tab @code{ omp_allocator_handle_t allocator)}
3022 @multitable @columnfractions .20 .80
3023 @item @emph{Prototype}: @tab @code{void* omp_aligned_alloc(size_t alignment,}
3024 @item @tab @code{ size_t size,}
3025 @item @tab @code{ omp_allocator_handle_t allocator=omp_null_allocator)}
3028 @item @emph{Fortran}:
3029 @multitable @columnfractions .20 .80
3030 @item @emph{Interface}: @tab @code{type(c_ptr) function omp_aligned_alloc(alignment, size, allocator) bind(C)}
3031 @item @tab @code{use, intrinsic :: iso_c_binding, only : c_ptr, c_size_t}
3032 @item @tab @code{integer (c_size_t), value :: alignment, size}
3033 @item @tab @code{integer (omp_allocator_handle_kind), value :: allocator}
3036 @item @emph{See also}:
3037 @ref{OMP_ALLOCATOR}, @ref{Memory allocation}, @ref{omp_set_default_allocator},
3038 @ref{omp_free}, @ref{omp_init_allocator}
3040 @item @emph{Reference}:
3041 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.13.6
3047 @subsection @code{omp_free} -- Freeing memory allocated with OpenMP routines
3049 @item @emph{Description}:
3050 The @code{omp_free} routine deallocates memory previously allocated by an
3051 OpenMP memory-management routine. The @var{ptr} argument must point to such
3052 memory or be a null pointer; if it is a null pointer, no operation is
3053 performed. If specified, the @var{allocator} argument must be either the
3054 memory allocator that was used for the allocation or @code{omp_null_allocator};
3055 if it is @code{omp_null_allocator}, the implementation will determine the value
3058 Calling @code{omp_free} invokes undefined behavior if the memory
3059 was already deallocated or when the used allocator has already been destroyed.
3062 @multitable @columnfractions .20 .80
3063 @item @emph{Prototype}: @tab @code{void omp_free(void *ptr,}
3064 @item @tab @code{ omp_allocator_handle_t allocator)}
3068 @multitable @columnfractions .20 .80
3069 @item @emph{Prototype}: @tab @code{void omp_free(void *ptr,}
3070 @item @tab @code{ omp_allocator_handle_t allocator=omp_null_allocator)}
3073 @item @emph{Fortran}:
3074 @multitable @columnfractions .20 .80
3075 @item @emph{Interface}: @tab @code{subroutine omp_free(ptr, allocator) bind(C)}
3076 @item @tab @code{use, intrinsic :: iso_c_binding, only : c_ptr}
3077 @item @tab @code{type (c_ptr), value :: ptr}
3078 @item @tab @code{integer (omp_allocator_handle_kind), value :: allocator}
3081 @item @emph{See also}:
3082 @ref{omp_alloc}, @ref{omp_aligned_alloc}, @ref{omp_calloc},
3083 @ref{omp_aligned_calloc}, @ref{omp_realloc}
3085 @item @emph{Reference}:
3086 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.7.7
3092 @subsection @code{omp_calloc} -- Allocate nullified memory with an allocator
3094 @item @emph{Description}:
3095 Allocate zero-initialized memory with the specified allocator, which can either
3096 be a predefined allocator, an allocator handle or @code{omp_null_allocator}. If
3097 the allocators is @code{omp_null_allocator}, the allocator specified by the
3098 @var{def-allocator-var} ICV is used. The to-be allocated memory is for an
3099 array with @var{nmemb} elements, each having a size of @var{size} bytes. Both
3100 @var{nmemb} and @var{size} must be nonnegative numbers; if either of them is
3101 zero, @code{omp_calloc} will return a null pointer. If successful, a pointer to
3102 the zero-initialized allocated memory is returned, otherwise the @code{fallback}
3103 trait of the allocator determines the behavior.
3105 In @code{target} regions, either the @code{dynamic_allocators} clause must
3106 appear on a @code{requires} directive in the same compilation unit -- or the
3107 @var{allocator} argument may only be a constant expression with the value of
3108 one of the predefined allocators and may not be @code{omp_null_allocator}.
3110 Memory allocated by @code{omp_calloc} must be freed using @code{omp_free}.
3113 @multitable @columnfractions .20 .80
3114 @item @emph{Prototype}: @tab @code{void* omp_calloc(size_t nmemb, size_t size,}
3115 @item @tab @code{ omp_allocator_handle_t allocator)}
3119 @multitable @columnfractions .20 .80
3120 @item @emph{Prototype}: @tab @code{void* omp_calloc(size_t nmemb, size_t size,}
3121 @item @tab @code{ omp_allocator_handle_t allocator=omp_null_allocator)}
3124 @item @emph{Fortran}:
3125 @multitable @columnfractions .20 .80
3126 @item @emph{Interface}: @tab @code{type(c_ptr) function omp_calloc(nmemb, size, allocator) bind(C)}
3127 @item @tab @code{use, intrinsic :: iso_c_binding, only : c_ptr, c_size_t}
3128 @item @tab @code{integer (c_size_t), value :: nmemb, size}
3129 @item @tab @code{integer (omp_allocator_handle_kind), value :: allocator}
3132 @item @emph{See also}:
3133 @ref{OMP_ALLOCATOR}, @ref{Memory allocation}, @ref{omp_set_default_allocator},
3134 @ref{omp_free}, @ref{omp_init_allocator}
3136 @item @emph{Reference}:
3137 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.13.8
3142 @node omp_aligned_calloc
3143 @subsection @code{omp_aligned_calloc} -- Allocate aligned nullified memory with an allocator
3145 @item @emph{Description}:
3146 Allocate zero-initialized memory with the specified allocator, which can either
3147 be a predefined allocator, an allocator handle or @code{omp_null_allocator}. If
3148 the allocators is @code{omp_null_allocator}, the allocator specified by the
3149 @var{def-allocator-var} ICV is used. The to-be allocated memory is for an
3150 array with @var{nmemb} elements, each having a size of @var{size} bytes. Both
3151 @var{nmemb} and @var{size} must be nonnegative numbers; if either of them is
3152 zero, @code{omp_aligned_calloc} will return a null pointer. @var{alignment}
3153 must be a positive power of two and @var{size} must be a multiple of the
3154 alignment; the alignment will be at least the maximal value required by
3155 @code{alignment} trait of the allocator and the value of the passed
3156 @var{alignment} argument. If successful, a pointer to the zero-initialized
3157 allocated memory is returned, otherwise the @code{fallback} trait of the
3158 allocator determines the behavior.
3160 In @code{target} regions, either the @code{dynamic_allocators} clause must
3161 appear on a @code{requires} directive in the same compilation unit -- or the
3162 @var{allocator} argument may only be a constant expression with the value of
3163 one of the predefined allocators and may not be @code{omp_null_allocator}.
3165 Memory allocated by @code{omp_aligned_calloc} must be freed using
3169 @multitable @columnfractions .20 .80
3170 @item @emph{Prototype}: @tab @code{void* omp_aligned_calloc(size_t nmemb, size_t size,}
3171 @item @tab @code{ omp_allocator_handle_t allocator)}
3175 @multitable @columnfractions .20 .80
3176 @item @emph{Prototype}: @tab @code{void* omp_aligned_calloc(size_t nmemb, size_t size,}
3177 @item @tab @code{ omp_allocator_handle_t allocator=omp_null_allocator)}
3180 @item @emph{Fortran}:
3181 @multitable @columnfractions .20 .80
3182 @item @emph{Interface}: @tab @code{type(c_ptr) function omp_aligned_calloc(nmemb, size, allocator) bind(C)}
3183 @item @tab @code{use, intrinsic :: iso_c_binding, only : c_ptr, c_size_t}
3184 @item @tab @code{integer (c_size_t), value :: nmemb, size}
3185 @item @tab @code{integer (omp_allocator_handle_kind), value :: allocator}
3188 @item @emph{See also}:
3189 @ref{OMP_ALLOCATOR}, @ref{Memory allocation}, @ref{omp_set_default_allocator},
3190 @ref{omp_free}, @ref{omp_init_allocator}
3192 @item @emph{Reference}:
3193 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.13.8
3199 @subsection @code{omp_realloc} -- Reallocate memory allocated with OpenMP routines
3201 @item @emph{Description}:
3202 The @code{omp_realloc} routine deallocates memory to which @var{ptr} points to
3203 and allocates new memory with the specified @var{allocator} argument; the
3204 new memory will have the content of the old memory up to the minimum of the
3205 old size and the new @var{size}, otherwise the content of the returned memory
3206 is unspecified. If the new allocator is the same as the old one, the routine
3207 tries to resize the existing memory allocation, returning the same address as
3208 @var{ptr} if successful. @var{ptr} must point to memory allocated by an OpenMP
3209 memory-management routine.
3211 The @var{allocator} and @var{free_allocator} arguments must be a predefined
3212 allocator, an allocator handle or @code{omp_null_allocator}. If
3213 @var{free_allocator} is @code{omp_null_allocator}, the implementation
3214 automatically determines the allocator used for the allocation of @var{ptr}.
3215 If @var{allocator} is @code{omp_null_allocator} and @var{ptr} is is not a
3216 null pointer, the same allocator as @code{free_allocator} is used and
3217 when @var{ptr} is a null pointer the allocator specified by the
3218 @var{def-allocator-var} ICV is used.
3220 The @var{size} must be a nonnegative number denoting the number of bytes to be
3221 allocated; if @var{size} is zero, @code{omp_realloc} will return free the
3222 memory and return a null pointer. When @var{size} is nonzero: if successful,
3223 a pointer to the allocated memory is returned, otherwise the @code{fallback}
3224 trait of the allocator determines the behavior.
3226 In @code{target} regions, either the @code{dynamic_allocators} clause must
3227 appear on a @code{requires} directive in the same compilation unit -- or the
3228 @var{free_allocator} and @var{allocator} arguments may only be a constant
3229 expression with the value of one of the predefined allocators and may not be
3230 @code{omp_null_allocator}.
3232 Memory allocated by @code{omp_realloc} must be freed using @code{omp_free}.
3233 Calling @code{omp_free} invokes undefined behavior if the memory
3234 was already deallocated or when the used allocator has already been destroyed.
3237 @multitable @columnfractions .20 .80
3238 @item @emph{Prototype}: @tab @code{void* omp_realloc(void *ptr, size_t size,}
3239 @item @tab @code{ omp_allocator_handle_t allocator,}
3240 @item @tab @code{ omp_allocator_handle_t free_allocator)}
3244 @multitable @columnfractions .20 .80
3245 @item @emph{Prototype}: @tab @code{void* omp_realloc(void *ptr, size_t size,}
3246 @item @tab @code{ omp_allocator_handle_t allocator=omp_null_allocator,}
3247 @item @tab @code{ omp_allocator_handle_t free_allocator=omp_null_allocator)}
3250 @item @emph{Fortran}:
3251 @multitable @columnfractions .20 .80
3252 @item @emph{Interface}: @tab @code{type(c_ptr) function omp_realloc(ptr, size, allocator, free_allocator) bind(C)}
3253 @item @tab @code{use, intrinsic :: iso_c_binding, only : c_ptr, c_size_t}
3254 @item @tab @code{type(C_ptr), value :: ptr}
3255 @item @tab @code{integer (c_size_t), value :: size}
3256 @item @tab @code{integer (omp_allocator_handle_kind), value :: allocator, free_allocator}
3259 @item @emph{See also}:
3260 @ref{OMP_ALLOCATOR}, @ref{Memory allocation}, @ref{omp_set_default_allocator},
3261 @ref{omp_free}, @ref{omp_init_allocator}
3263 @item @emph{Reference}:
3264 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.7.9
3269 @c @node Tool Control Routine
3270 @c @section Tool Control Routine
3274 @node Environment Display Routine
3275 @section Environment Display Routine
3277 Routine to display the OpenMP version number and the initial value of ICVs.
3278 It has C linkage and does not throw exceptions.
3281 * omp_display_env:: print the initial ICV values
3284 @node omp_display_env
3285 @subsection @code{omp_display_env} -- print the initial ICV values
3287 @item @emph{Description}:
3288 Each time this routine is invoked, the OpenMP version number and initial value
3289 of internal control variables (ICVs) is printed on @code{stderr}. The displayed
3290 values are those at startup after evaluating the environment variables; later
3291 calls to API routines or clauses used in enclosing constructs do not affect
3294 If the @var{verbose} argument is @code{false}, only the OpenMP version and
3295 standard OpenMP ICVs are shown; if it is @code{true}, additionally, the
3296 GCC-specific ICVs are shown.
3298 The output consists of multiple lines and starts with
3299 @samp{OPENMP DISPLAY ENVIRONMENT BEGIN} followed by the name-value lines and
3300 ends with @samp{OPENMP DISPLAY ENVIRONMENT END}. The @var{name} is followed by
3301 an equal sign and the @var{value} is enclosed in single quotes.
3303 The first line has as @var{name} either @samp{_OPENMP} or @samp{openmp_version}
3304 and shows as value the supported OpenMP version number (4-digit year, 2-digit
3305 month) of the implementation, matching the value of the @code{_OPENMP} macro
3306 and, in Fortran, the named constant @code{openmp_version}.
3308 In each of the succeeding lines, the @var{name} matches the environment-variable
3309 name of an ICV and shows its value. Those line are might be prefixed by pair of
3310 brackets and a space, where the brackets enclose a comma-separated list of
3311 devices to which the ICV-value combination applies to; the value can either be a
3312 numeric device number or an abstract name denoting all devices (@code{all}), the
3313 initial host device (@code{host}) or all devices but the host (@code{device}).
3314 Note that the same ICV might be printed multiple times for multiple devices,
3315 even if all have the same value.
3317 The effect when invoked from within a @code{target} region is unspecified.
3320 @multitable @columnfractions .20 .80
3321 @item @emph{Prototype}: @tab @code{void omp_display_env(int verbose)}
3324 @item @emph{Fortran}:
3325 @multitable @columnfractions .20 .80
3326 @item @emph{Interface}: @tab @code{subroutine omp_display_env(vebose)}
3327 @item @tab @code{logical, intent(in) :: verbose}
3330 @item @emph{Example}:
3331 Note that the GCC-specific ICVs, such as the shown @code{GOMP_SPINCOUNT},
3332 are only printed when @var{varbose} set to @code{true}.
3335 OPENMP DISPLAY ENVIRONMENT BEGIN
3337 [host] OMP_DYNAMIC = 'FALSE'
3338 [host] OMP_NESTED = 'FALSE'
3339 [all] OMP_CANCELLATION = 'FALSE'
3341 [host] GOMP_SPINCOUNT = '300000'
3342 OPENMP DISPLAY ENVIRONMENT END
3346 @item @emph{See also}:
3347 @ref{OMP_DISPLAY_ENV}, @ref{Environment Variables},
3348 @ref{Implementation-defined ICV Initialization}
3350 @item @emph{Reference}:
3351 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 3.15
3355 @c ---------------------------------------------------------------------
3356 @c OpenMP Environment Variables
3357 @c ---------------------------------------------------------------------
3359 @node Environment Variables
3360 @chapter OpenMP Environment Variables
3362 The environment variables which beginning with @env{OMP_} are defined by
3363 section 4 of the OpenMP specification in version 4.5 or in a later version
3364 of the specification, while those beginning with @env{GOMP_} are GNU extensions.
3365 Most @env{OMP_} environment variables have an associated internal control
3368 For any OpenMP environment variable that sets an ICV and is neither
3369 @code{OMP_DEFAULT_DEVICE} nor has global ICV scope, associated
3370 device-specific environment variables exist. For them, the environment
3371 variable without suffix affects the host. The suffix @code{_DEV_} followed
3372 by a non-negative device number less that the number of available devices sets
3373 the ICV for the corresponding device. The suffix @code{_DEV} sets the ICV
3374 of all non-host devices for which a device-specific corresponding environment
3375 variable has not been set while the @code{_ALL} suffix sets the ICV of all
3376 host and non-host devices for which a more specific corresponding environment
3377 variable is not set.
3380 * OMP_ALLOCATOR:: Set the default allocator
3381 * OMP_AFFINITY_FORMAT:: Set the format string used for affinity display
3382 * OMP_CANCELLATION:: Set whether cancellation is activated
3383 * OMP_DISPLAY_AFFINITY:: Display thread affinity information
3384 * OMP_DISPLAY_ENV:: Show OpenMP version and environment variables
3385 * OMP_DEFAULT_DEVICE:: Set the device used in target regions
3386 * OMP_DYNAMIC:: Dynamic adjustment of threads
3387 * OMP_MAX_ACTIVE_LEVELS:: Set the maximum number of nested parallel regions
3388 * OMP_MAX_TASK_PRIORITY:: Set the maximum task priority value
3389 * OMP_NESTED:: Nested parallel regions
3390 * OMP_NUM_TEAMS:: Specifies the number of teams to use by teams region
3391 * OMP_NUM_THREADS:: Specifies the number of threads to use
3392 * OMP_PROC_BIND:: Whether threads may be moved between CPUs
3393 * OMP_PLACES:: Specifies on which CPUs the threads should be placed
3394 * OMP_STACKSIZE:: Set default thread stack size
3395 * OMP_SCHEDULE:: How threads are scheduled
3396 * OMP_TARGET_OFFLOAD:: Controls offloading behavior
3397 * OMP_TEAMS_THREAD_LIMIT:: Set the maximum number of threads imposed by teams
3398 * OMP_THREAD_LIMIT:: Set the maximum number of threads
3399 * OMP_WAIT_POLICY:: How waiting threads are handled
3400 * GOMP_CPU_AFFINITY:: Bind threads to specific CPUs
3401 * GOMP_DEBUG:: Enable debugging output
3402 * GOMP_STACKSIZE:: Set default thread stack size
3403 * GOMP_SPINCOUNT:: Set the busy-wait spin count
3404 * GOMP_RTEMS_THREAD_POOLS:: Set the RTEMS specific thread pools
3409 @section @env{OMP_ALLOCATOR} -- Set the default allocator
3410 @cindex Environment Variable
3412 @item @emph{ICV:} @var{def-allocator-var}
3413 @item @emph{Scope:} data environment
3414 @item @emph{Description}:
3415 Sets the default allocator that is used when no allocator has been specified
3416 in the @code{allocate} or @code{allocator} clause or if an OpenMP memory
3417 routine is invoked with the @code{omp_null_allocator} allocator.
3418 If unset, @code{omp_default_mem_alloc} is used.
3420 The value can either be a predefined allocator or a predefined memory space
3421 or a predefined memory space followed by a colon and a comma-separated list
3422 of memory trait and value pairs, separated by @code{=}.
3424 Note: The corresponding device environment variables are currently not
3425 supported. Therefore, the non-host @var{def-allocator-var} ICVs are always
3426 initialized to @code{omp_default_mem_alloc}. However, on all devices,
3427 the @code{omp_set_default_allocator} API routine can be used to change
3430 @multitable @columnfractions .45 .45
3431 @headitem Predefined allocators @tab Associated predefined memory spaces
3432 @item omp_default_mem_alloc @tab omp_default_mem_space
3433 @item omp_large_cap_mem_alloc @tab omp_large_cap_mem_space
3434 @item omp_const_mem_alloc @tab omp_const_mem_space
3435 @item omp_high_bw_mem_alloc @tab omp_high_bw_mem_space
3436 @item omp_low_lat_mem_alloc @tab omp_low_lat_mem_space
3437 @item omp_cgroup_mem_alloc @tab omp_low_lat_mem_space (implementation defined)
3438 @item omp_pteam_mem_alloc @tab omp_low_lat_mem_space (implementation defined)
3439 @item omp_thread_mem_alloc @tab omp_low_lat_mem_space (implementation defined)
3442 The predefined allocators use the default values for the traits,
3443 as listed below. Except that the last three allocators have the
3444 @code{access} trait set to @code{cgroup}, @code{pteam}, and
3445 @code{thread}, respectively.
3447 @multitable @columnfractions .25 .40 .25
3448 @headitem Trait @tab Allowed values @tab Default value
3449 @item @code{sync_hint} @tab @code{contended}, @code{uncontended},
3450 @code{serialized}, @code{private}
3451 @tab @code{contended}
3452 @item @code{alignment} @tab Positive integer being a power of two
3454 @item @code{access} @tab @code{all}, @code{cgroup},
3455 @code{pteam}, @code{thread}
3457 @item @code{pool_size} @tab Positive integer
3458 @tab See @ref{Memory allocation}
3459 @item @code{fallback} @tab @code{default_mem_fb}, @code{null_fb},
3460 @code{abort_fb}, @code{allocator_fb}
3462 @item @code{fb_data} @tab @emph{unsupported as it needs an allocator handle}
3464 @item @code{pinned} @tab @code{true}, @code{false}
3466 @item @code{partition} @tab @code{environment}, @code{nearest},
3467 @code{blocked}, @code{interleaved}
3468 @tab @code{environment}
3471 For the @code{fallback} trait, the default value is @code{null_fb} for the
3472 @code{omp_default_mem_alloc} allocator and any allocator that is associated
3473 with device memory; for all other other allocators, it is @code{default_mem_fb}
3478 OMP_ALLOCATOR=omp_high_bw_mem_alloc
3479 OMP_ALLOCATOR=omp_large_cap_mem_space
3480 OMP_ALLOCATOR=omp_low_lat_mem_space:pinned=true,partition=nearest
3483 @item @emph{See also}:
3484 @ref{Memory allocation}, @ref{omp_get_default_allocator},
3485 @ref{omp_set_default_allocator}, @ref{Offload-Target Specifics}
3487 @item @emph{Reference}:
3488 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 6.21
3493 @node OMP_AFFINITY_FORMAT
3494 @section @env{OMP_AFFINITY_FORMAT} -- Set the format string used for affinity display
3495 @cindex Environment Variable
3497 @item @emph{ICV:} @var{affinity-format-var}
3498 @item @emph{Scope:} device
3499 @item @emph{Description}:
3500 Sets the format string used when displaying OpenMP thread affinity information.
3501 Special values are output using @code{%} followed by an optional size
3502 specification and then either the single-character field type or its long
3503 name enclosed in curly braces; using @code{%%} displays a literal percent.
3504 The size specification consists of an optional @code{0.} or @code{.} followed
3505 by a positive integer, specifying the minimal width of the output. With
3506 @code{0.} and numerical values, the output is padded with zeros on the left;
3507 with @code{.}, the output is padded by spaces on the left; otherwise, the
3508 output is padded by spaces on the right. If unset, the value is
3509 ``@code{level %L thread %i affinity %A}''.
3511 Supported field types are:
3513 @multitable @columnfractions .10 .25 .60
3514 @item t @tab team_num @tab value returned by @code{omp_get_team_num}
3515 @item T @tab num_teams @tab value returned by @code{omp_get_num_teams}
3516 @item L @tab nesting_level @tab value returned by @code{omp_get_level}
3517 @item n @tab thread_num @tab value returned by @code{omp_get_thread_num}
3518 @item N @tab num_threads @tab value returned by @code{omp_get_num_threads}
3519 @item a @tab ancestor_tnum
3520 @tab value returned by
3521 @code{omp_get_ancestor_thread_num(omp_get_level()-1)}
3522 @item H @tab host @tab name of the host that executes the thread
3523 @item P @tab process_id @tab process identifier
3524 @item i @tab native_thread_id @tab native thread identifier
3525 @item A @tab thread_affinity
3526 @tab comma separated list of integer values or ranges, representing the
3527 processors on which a process might execute, subject to affinity
3531 For instance, after setting
3534 OMP_AFFINITY_FORMAT="%0.2a!%n!%.4L!%N;%.2t;%0.2T;%@{team_num@};%@{num_teams@};%A"
3537 with either @code{OMP_DISPLAY_AFFINITY} being set or when calling
3538 @code{omp_display_affinity} with @code{NULL} or an empty string, the program
3539 might display the following:
3542 00!0! 1!4; 0;01;0;1;0-11
3543 00!3! 1!4; 0;01;0;1;0-11
3544 00!2! 1!4; 0;01;0;1;0-11
3545 00!1! 1!4; 0;01;0;1;0-11
3548 @item @emph{See also}:
3549 @ref{OMP_DISPLAY_AFFINITY}
3551 @item @emph{Reference}:
3552 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 6.14
3557 @node OMP_CANCELLATION
3558 @section @env{OMP_CANCELLATION} -- Set whether cancellation is activated
3559 @cindex Environment Variable
3561 @item @emph{ICV:} @var{cancel-var}
3562 @item @emph{Scope:} global
3563 @item @emph{Description}:
3564 If set to @code{TRUE}, the cancellation is activated. If set to @code{FALSE} or
3565 if unset, cancellation is disabled and the @code{cancel} construct is ignored.
3567 @item @emph{See also}:
3568 @ref{omp_get_cancellation}
3570 @item @emph{Reference}:
3571 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.11
3576 @node OMP_DISPLAY_AFFINITY
3577 @section @env{OMP_DISPLAY_AFFINITY} -- Display thread affinity information
3578 @cindex Environment Variable
3580 @item @emph{ICV:} @var{display-affinity-var}
3581 @item @emph{Scope:} global
3582 @item @emph{Description}:
3583 If set to @code{FALSE} or if unset, affinity displaying is disabled.
3584 If set to @code{TRUE}, the runtime displays affinity information about
3585 OpenMP threads in a parallel region upon entering the region and every time
3588 @item @emph{See also}:
3589 @ref{OMP_AFFINITY_FORMAT}
3591 @item @emph{Reference}:
3592 @uref{https://www.openmp.org, OpenMP specification v5.0}, Section 6.13
3598 @node OMP_DISPLAY_ENV
3599 @section @env{OMP_DISPLAY_ENV} -- Show OpenMP version and environment variables
3600 @cindex Environment Variable
3602 @item @emph{ICV:} none
3603 @item @emph{Scope:} not applicable
3604 @item @emph{Description}:
3605 If set to @code{TRUE}, the runtime displays the same information to
3606 @code{stderr} as shown by the @code{omp_display_env} routine invoked with
3607 @var{verbose} argument set to @code{false}. If set to @code{VERBOSE}, the same
3608 information is shown as invoking the routine with @var{verbose} set to
3609 @code{true}. If unset or set to @code{FALSE}, this information is not shown.
3610 The result for any other value is unspecified.
3612 @item @emph{See also}:
3613 @ref{omp_display_env}
3615 @item @emph{Reference}:
3616 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.12
3621 @node OMP_DEFAULT_DEVICE
3622 @section @env{OMP_DEFAULT_DEVICE} -- Set the device used in target regions
3623 @cindex Environment Variable
3625 @item @emph{ICV:} @var{default-device-var}
3626 @item @emph{Scope:} data environment
3627 @item @emph{Description}:
3628 Set to choose the device which is used in a @code{target} region, unless the
3629 value is overridden by @code{omp_set_default_device} or by a @code{device}
3630 clause. The value shall be the nonnegative device number. If no device with
3631 the given device number exists, the code is executed on the host. If unset,
3632 @env{OMP_TARGET_OFFLOAD} is @code{mandatory} and no non-host devices are
3633 available, it is set to @code{omp_invalid_device}. Otherwise, if unset,
3634 device number 0 is used.
3637 @item @emph{See also}:
3638 @ref{omp_get_default_device}, @ref{omp_set_default_device},
3639 @ref{OMP_TARGET_OFFLOAD}
3641 @item @emph{Reference}:
3642 @uref{https://www.openmp.org, OpenMP specification v5.2}, Section 21.2.7
3648 @section @env{OMP_DYNAMIC} -- Dynamic adjustment of threads
3649 @cindex Environment Variable
3651 @item @emph{ICV:} @var{dyn-var}
3652 @item @emph{Scope:} global
3653 @item @emph{Description}:
3654 Enable or disable the dynamic adjustment of the number of threads
3655 within a team. The value of this environment variable shall be
3656 @code{TRUE} or @code{FALSE}. If undefined, dynamic adjustment is
3657 disabled by default.
3659 @item @emph{See also}:
3660 @ref{omp_set_dynamic}
3662 @item @emph{Reference}:
3663 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.3
3668 @node OMP_MAX_ACTIVE_LEVELS
3669 @section @env{OMP_MAX_ACTIVE_LEVELS} -- Set the maximum number of nested parallel regions
3670 @cindex Environment Variable
3672 @item @emph{ICV:} @var{max-active-levels-var}
3673 @item @emph{Scope:} data environment
3674 @item @emph{Description}:
3675 Specifies the initial value for the maximum number of nested parallel
3676 regions. The value of this variable shall be a positive integer.
3677 If undefined, then if @env{OMP_NESTED} is defined and set to true, or
3678 if @env{OMP_NUM_THREADS} or @env{OMP_PROC_BIND} are defined and set to
3679 a list with more than one item, the maximum number of nested parallel
3680 regions is initialized to the largest number supported, otherwise
3683 @item @emph{See also}:
3684 @ref{omp_set_max_active_levels}, @ref{OMP_NESTED}, @ref{OMP_PROC_BIND},
3685 @ref{OMP_NUM_THREADS}
3688 @item @emph{Reference}:
3689 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.9
3694 @node OMP_MAX_TASK_PRIORITY
3695 @section @env{OMP_MAX_TASK_PRIORITY} -- Set the maximum priority
3696 number that can be set for a task.
3697 @cindex Environment Variable
3699 @item @emph{ICV:} @var{max-task-priority-var}
3700 @item @emph{Scope:} global
3701 @item @emph{Description}:
3702 Specifies the initial value for the maximum priority value that can be
3703 set for a task. The value of this variable shall be a non-negative
3704 integer, and zero is allowed. If undefined, the default priority is
3707 @item @emph{See also}:
3708 @ref{omp_get_max_task_priority}
3710 @item @emph{Reference}:
3711 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.14
3717 @section @env{OMP_NESTED} -- Nested parallel regions
3718 @cindex Environment Variable
3719 @cindex Implementation specific setting
3721 @item @emph{ICV:} @var{max-active-levels-var}
3722 @item @emph{Scope:} data environment
3723 @item @emph{Description}:
3724 Enable or disable nested parallel regions, i.e., whether team members
3725 are allowed to create new teams. The value of this environment variable
3726 shall be @code{TRUE} or @code{FALSE}. If set to @code{TRUE}, the number
3727 of maximum active nested regions supported is by default set to the
3728 maximum supported, otherwise it is set to one. If
3729 @env{OMP_MAX_ACTIVE_LEVELS} is defined, its setting overrides this
3730 setting. If both are undefined, nested parallel regions are enabled if
3731 @env{OMP_NUM_THREADS} or @env{OMP_PROC_BINDS} are defined to a list with
3732 more than one item, otherwise they are disabled by default.
3734 Note that the @code{OMP_NESTED} environment variable was deprecated in
3735 the OpenMP specification 5.2 in favor of @code{OMP_MAX_ACTIVE_LEVELS}.
3737 @item @emph{See also}:
3738 @ref{omp_set_max_active_levels}, @ref{omp_set_nested},
3739 @ref{OMP_MAX_ACTIVE_LEVELS}
3741 @item @emph{Reference}:
3742 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.6
3748 @section @env{OMP_NUM_TEAMS} -- Specifies the number of teams to use by teams region
3749 @cindex Environment Variable
3751 @item @emph{ICV:} @var{nteams-var}
3752 @item @emph{Scope:} device
3753 @item @emph{Description}:
3754 Specifies the upper bound for number of teams to use in teams regions
3755 without explicit @code{num_teams} clause. The value of this variable shall
3756 be a positive integer. If undefined it defaults to 0 which means
3757 implementation defined upper bound.
3759 @item @emph{See also}:
3760 @ref{omp_set_num_teams}
3762 @item @emph{Reference}:
3763 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 6.23
3768 @node OMP_NUM_THREADS
3769 @section @env{OMP_NUM_THREADS} -- Specifies the number of threads to use
3770 @cindex Environment Variable
3771 @cindex Implementation specific setting
3773 @item @emph{ICV:} @var{nthreads-var}
3774 @item @emph{Scope:} data environment
3775 @item @emph{Description}:
3776 Specifies the default number of threads to use in parallel regions. The
3777 value of this variable shall be a comma-separated list of positive integers;
3778 the value specifies the number of threads to use for the corresponding nested
3779 level. Specifying more than one item in the list automatically enables
3780 nesting by default. If undefined one thread per CPU is used.
3782 When a list with more than value is specified, it also affects the
3783 @var{max-active-levels-var} ICV as described in @ref{OMP_MAX_ACTIVE_LEVELS}.
3785 @item @emph{See also}:
3786 @ref{omp_set_num_threads}, @ref{OMP_MAX_ACTIVE_LEVELS}
3788 @item @emph{Reference}:
3789 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.2
3795 @section @env{OMP_PROC_BIND} -- Whether threads may be moved between CPUs
3796 @cindex Environment Variable
3798 @item @emph{ICV:} @var{bind-var}
3799 @item @emph{Scope:} data environment
3800 @item @emph{Description}:
3801 Specifies whether threads may be moved between processors. If set to
3802 @code{TRUE}, OpenMP threads should not be moved; if set to @code{FALSE}
3803 they may be moved. Alternatively, a comma separated list with the
3804 values @code{PRIMARY}, @code{MASTER}, @code{CLOSE} and @code{SPREAD} can
3805 be used to specify the thread affinity policy for the corresponding nesting
3806 level. With @code{PRIMARY} and @code{MASTER} the worker threads are in the
3807 same place partition as the primary thread. With @code{CLOSE} those are
3808 kept close to the primary thread in contiguous place partitions. And
3809 with @code{SPREAD} a sparse distribution
3810 across the place partitions is used. Specifying more than one item in the
3811 list automatically enables nesting by default.
3813 When a list is specified, it also affects the @var{max-active-levels-var} ICV
3814 as described in @ref{OMP_MAX_ACTIVE_LEVELS}.
3816 When undefined, @env{OMP_PROC_BIND} defaults to @code{TRUE} when
3817 @env{OMP_PLACES} or @env{GOMP_CPU_AFFINITY} is set and @code{FALSE} otherwise.
3819 @item @emph{See also}:
3820 @ref{omp_get_proc_bind}, @ref{GOMP_CPU_AFFINITY}, @ref{OMP_PLACES},
3821 @ref{OMP_MAX_ACTIVE_LEVELS}
3823 @item @emph{Reference}:
3824 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.4
3830 @section @env{OMP_PLACES} -- Specifies on which CPUs the threads should be placed
3831 @cindex Environment Variable
3833 @item @emph{ICV:} @var{place-partition-var}
3834 @item @emph{Scope:} implicit tasks
3835 @item @emph{Description}:
3836 The thread placement can be either specified using an abstract name or by an
3837 explicit list of the places. The abstract names @code{threads}, @code{cores},
3838 @code{sockets}, @code{ll_caches} and @code{numa_domains} can be optionally
3839 followed by a positive number in parentheses, which denotes the how many places
3840 shall be created. With @code{threads} each place corresponds to a single
3841 hardware thread; @code{cores} to a single core with the corresponding number of
3842 hardware threads; with @code{sockets} the place corresponds to a single
3843 socket; with @code{ll_caches} to a set of cores that shares the last level
3844 cache on the device; and @code{numa_domains} to a set of cores for which their
3845 closest memory on the device is the same memory and at a similar distance from
3846 the cores. The resulting placement can be shown by setting the
3847 @env{OMP_DISPLAY_ENV} environment variable.
3849 Alternatively, the placement can be specified explicitly as comma-separated
3850 list of places. A place is specified by set of nonnegative numbers in curly
3851 braces, denoting the hardware threads. The curly braces can be omitted
3852 when only a single number has been specified. The hardware threads
3853 belonging to a place can either be specified as comma-separated list of
3854 nonnegative thread numbers or using an interval. Multiple places can also be
3855 either specified by a comma-separated list of places or by an interval. To
3856 specify an interval, a colon followed by the count is placed after
3857 the hardware thread number or the place. Optionally, the length can be
3858 followed by a colon and the stride number -- otherwise a unit stride is
3859 assumed. Placing an exclamation mark (@code{!}) directly before a curly
3860 brace or numbers inside the curly braces (excluding intervals)
3861 excludes those hardware threads.
3863 For instance, the following specifies the same places list:
3864 @code{"@{0,1,2@}, @{3,4,6@}, @{7,8,9@}, @{10,11,12@}"};
3865 @code{"@{0:3@}, @{3:3@}, @{7:3@}, @{10:3@}"}; and @code{"@{0:2@}:4:3"}.
3867 If @env{OMP_PLACES} and @env{GOMP_CPU_AFFINITY} are unset and
3868 @env{OMP_PROC_BIND} is either unset or @code{false}, threads may be moved
3869 between CPUs following no placement policy.
3871 @item @emph{See also}:
3872 @ref{OMP_PROC_BIND}, @ref{GOMP_CPU_AFFINITY}, @ref{omp_get_proc_bind},
3873 @ref{OMP_DISPLAY_ENV}
3875 @item @emph{Reference}:
3876 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.5
3882 @section @env{OMP_STACKSIZE} -- Set default thread stack size
3883 @cindex Environment Variable
3885 @item @emph{ICV:} @var{stacksize-var}
3886 @item @emph{Scope:} device
3887 @item @emph{Description}:
3888 Set the default thread stack size in kilobytes, unless the number
3889 is suffixed by @code{B}, @code{K}, @code{M} or @code{G}, in which
3890 case the size is, respectively, in bytes, kilobytes, megabytes
3891 or gigabytes. This is different from @code{pthread_attr_setstacksize}
3892 which gets the number of bytes as an argument. If the stack size cannot
3893 be set due to system constraints, an error is reported and the initial
3894 stack size is left unchanged. If undefined, the stack size is system
3897 @item @emph{See also}:
3898 @ref{GOMP_STACKSIZE}
3900 @item @emph{Reference}:
3901 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.7
3907 @section @env{OMP_SCHEDULE} -- How threads are scheduled
3908 @cindex Environment Variable
3909 @cindex Implementation specific setting
3911 @item @emph{ICV:} @var{run-sched-var}
3912 @item @emph{Scope:} data environment
3913 @item @emph{Description}:
3914 Allows to specify @code{schedule type} and @code{chunk size}.
3915 The value of the variable shall have the form: @code{type[,chunk]} where
3916 @code{type} is one of @code{static}, @code{dynamic}, @code{guided} or @code{auto}
3917 The optional @code{chunk} size shall be a positive integer. If undefined,
3918 dynamic scheduling and a chunk size of 1 is used.
3920 @item @emph{See also}:
3921 @ref{omp_set_schedule}
3923 @item @emph{Reference}:
3924 @uref{https://www.openmp.org, OpenMP specification v4.5}, Sections 2.7.1.1 and 4.1
3929 @node OMP_TARGET_OFFLOAD
3930 @section @env{OMP_TARGET_OFFLOAD} -- Controls offloading behavior
3931 @cindex Environment Variable
3932 @cindex Implementation specific setting
3934 @item @emph{ICV:} @var{target-offload-var}
3935 @item @emph{Scope:} global
3936 @item @emph{Description}:
3937 Specifies the behavior with regard to offloading code to a device. This
3938 variable can be set to one of three values - @code{MANDATORY}, @code{DISABLED}
3941 If set to @code{MANDATORY}, the program terminates with an error if
3942 any device construct or device memory routine uses a device that is unavailable
3943 or not supported by the implementation, or uses a non-conforming device number.
3944 If set to @code{DISABLED}, then offloading is disabled and all code runs on
3945 the host. If set to @code{DEFAULT}, the program tries offloading to the
3946 device first, then falls back to running code on the host if it cannot.
3948 If undefined, then the program behaves as if @code{DEFAULT} was set.
3950 Note: Even with @code{MANDATORY}, no run-time termination is performed when
3951 the device number in a @code{device} clause or argument to a device memory
3952 routine is for host, which includes using the device number in the
3953 @var{default-device-var} ICV. However, the initial value of
3954 the @var{default-device-var} ICV is affected by @code{MANDATORY}.
3956 @item @emph{See also}:
3957 @ref{OMP_DEFAULT_DEVICE}
3959 @item @emph{Reference}:
3960 @uref{https://www.openmp.org, OpenMP specification v5.2}, Section 21.2.8
3965 @node OMP_TEAMS_THREAD_LIMIT
3966 @section @env{OMP_TEAMS_THREAD_LIMIT} -- Set the maximum number of threads imposed by teams
3967 @cindex Environment Variable
3969 @item @emph{ICV:} @var{teams-thread-limit-var}
3970 @item @emph{Scope:} device
3971 @item @emph{Description}:
3972 Specifies an upper bound for the number of threads to use by each contention
3973 group created by a teams construct without explicit @code{thread_limit}
3974 clause. The value of this variable shall be a positive integer. If undefined,
3975 the value of 0 is used which stands for an implementation defined upper
3978 @item @emph{See also}:
3979 @ref{OMP_THREAD_LIMIT}, @ref{omp_set_teams_thread_limit}
3981 @item @emph{Reference}:
3982 @uref{https://www.openmp.org, OpenMP specification v5.1}, Section 6.24
3987 @node OMP_THREAD_LIMIT
3988 @section @env{OMP_THREAD_LIMIT} -- Set the maximum number of threads
3989 @cindex Environment Variable
3991 @item @emph{ICV:} @var{thread-limit-var}
3992 @item @emph{Scope:} data environment
3993 @item @emph{Description}:
3994 Specifies the number of threads to use for the whole program. The
3995 value of this variable shall be a positive integer. If undefined,
3996 the number of threads is not limited.
3998 @item @emph{See also}:
3999 @ref{OMP_NUM_THREADS}, @ref{omp_get_thread_limit}
4001 @item @emph{Reference}:
4002 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.10
4007 @node OMP_WAIT_POLICY
4008 @section @env{OMP_WAIT_POLICY} -- How waiting threads are handled
4009 @cindex Environment Variable
4011 @item @emph{Description}:
4012 Specifies whether waiting threads should be active or passive. If
4013 the value is @code{PASSIVE}, waiting threads should not consume CPU
4014 power while waiting; while the value is @code{ACTIVE} specifies that
4015 they should. If undefined, threads wait actively for a short time
4016 before waiting passively.
4018 @item @emph{See also}:
4019 @ref{GOMP_SPINCOUNT}
4021 @item @emph{Reference}:
4022 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.8
4027 @node GOMP_CPU_AFFINITY
4028 @section @env{GOMP_CPU_AFFINITY} -- Bind threads to specific CPUs
4029 @cindex Environment Variable
4031 @item @emph{Description}:
4032 Binds threads to specific CPUs. The variable should contain a space-separated
4033 or comma-separated list of CPUs. This list may contain different kinds of
4034 entries: either single CPU numbers in any order, a range of CPUs (M-N)
4035 or a range with some stride (M-N:S). CPU numbers are zero based. For example,
4036 @code{GOMP_CPU_AFFINITY="0 3 1-2 4-15:2"} binds the initial thread
4037 to CPU 0, the second to CPU 3, the third to CPU 1, the fourth to
4038 CPU 2, the fifth to CPU 4, the sixth through tenth to CPUs 6, 8, 10, 12,
4039 and 14 respectively and then starts assigning back from the beginning of
4040 the list. @code{GOMP_CPU_AFFINITY=0} binds all threads to CPU 0.
4042 There is no libgomp library routine to determine whether a CPU affinity
4043 specification is in effect. As a workaround, language-specific library
4044 functions, e.g., @code{getenv} in C or @code{GET_ENVIRONMENT_VARIABLE} in
4045 Fortran, may be used to query the setting of the @code{GOMP_CPU_AFFINITY}
4046 environment variable. A defined CPU affinity on startup cannot be changed
4047 or disabled during the runtime of the application.
4049 If both @env{GOMP_CPU_AFFINITY} and @env{OMP_PROC_BIND} are set,
4050 @env{OMP_PROC_BIND} has a higher precedence. If neither has been set and
4051 @env{OMP_PROC_BIND} is unset, or when @env{OMP_PROC_BIND} is set to
4052 @code{FALSE}, the host system handles the assignment of threads to CPUs.
4054 @item @emph{See also}:
4055 @ref{OMP_PLACES}, @ref{OMP_PROC_BIND}
4061 @section @env{GOMP_DEBUG} -- Enable debugging output
4062 @cindex Environment Variable
4064 @item @emph{Description}:
4065 Enable debugging output. The variable should be set to @code{0}
4066 (disabled, also the default if not set), or @code{1} (enabled).
4068 If enabled, some debugging output is printed during execution.
4069 This is currently not specified in more detail, and subject to change.
4074 @node GOMP_STACKSIZE
4075 @section @env{GOMP_STACKSIZE} -- Set default thread stack size
4076 @cindex Environment Variable
4077 @cindex Implementation specific setting
4079 @item @emph{Description}:
4080 Set the default thread stack size in kilobytes. This is different from
4081 @code{pthread_attr_setstacksize} which gets the number of bytes as an
4082 argument. If the stack size cannot be set due to system constraints, an
4083 error is reported and the initial stack size is left unchanged. If undefined,
4084 the stack size is system dependent.
4086 @item @emph{See also}:
4089 @item @emph{Reference}:
4090 @uref{https://gcc.gnu.org/ml/gcc-patches/2006-06/msg00493.html,
4091 GCC Patches Mailinglist},
4092 @uref{https://gcc.gnu.org/ml/gcc-patches/2006-06/msg00496.html,
4093 GCC Patches Mailinglist}
4098 @node GOMP_SPINCOUNT
4099 @section @env{GOMP_SPINCOUNT} -- Set the busy-wait spin count
4100 @cindex Environment Variable
4101 @cindex Implementation specific setting
4103 @item @emph{Description}:
4104 Determines how long a threads waits actively with consuming CPU power
4105 before waiting passively without consuming CPU power. The value may be
4106 either @code{INFINITE}, @code{INFINITY} to always wait actively or an
4107 integer which gives the number of spins of the busy-wait loop. The
4108 integer may optionally be followed by the following suffixes acting
4109 as multiplication factors: @code{k} (kilo, thousand), @code{M} (mega,
4110 million), @code{G} (giga, billion), or @code{T} (tera, trillion).
4111 If undefined, 0 is used when @env{OMP_WAIT_POLICY} is @code{PASSIVE},
4112 300,000 is used when @env{OMP_WAIT_POLICY} is undefined and
4113 30 billion is used when @env{OMP_WAIT_POLICY} is @code{ACTIVE}.
4114 If there are more OpenMP threads than available CPUs, 1000 and 100
4115 spins are used for @env{OMP_WAIT_POLICY} being @code{ACTIVE} or
4116 undefined, respectively; unless the @env{GOMP_SPINCOUNT} is lower
4117 or @env{OMP_WAIT_POLICY} is @code{PASSIVE}.
4119 @item @emph{See also}:
4120 @ref{OMP_WAIT_POLICY}
4125 @node GOMP_RTEMS_THREAD_POOLS
4126 @section @env{GOMP_RTEMS_THREAD_POOLS} -- Set the RTEMS specific thread pools
4127 @cindex Environment Variable
4128 @cindex Implementation specific setting
4130 @item @emph{Description}:
4131 This environment variable is only used on the RTEMS real-time operating system.
4132 It determines the scheduler instance specific thread pools. The format for
4133 @env{GOMP_RTEMS_THREAD_POOLS} is a list of optional
4134 @code{<thread-pool-count>[$<priority>]@@<scheduler-name>} configurations
4135 separated by @code{:} where:
4137 @item @code{<thread-pool-count>} is the thread pool count for this scheduler
4139 @item @code{$<priority>} is an optional priority for the worker threads of a
4140 thread pool according to @code{pthread_setschedparam}. In case a priority
4141 value is omitted, then a worker thread inherits the priority of the OpenMP
4142 primary thread that created it. The priority of the worker thread is not
4143 changed after creation, even if a new OpenMP primary thread using the worker has
4144 a different priority.
4145 @item @code{@@<scheduler-name>} is the scheduler instance name according to the
4146 RTEMS application configuration.
4148 In case no thread pool configuration is specified for a scheduler instance,
4149 then each OpenMP primary thread of this scheduler instance uses its own
4150 dynamically allocated thread pool. To limit the worker thread count of the
4151 thread pools, each OpenMP primary thread must call @code{omp_set_num_threads}.
4152 @item @emph{Example}:
4153 Lets suppose we have three scheduler instances @code{IO}, @code{WRK0}, and
4154 @code{WRK1} with @env{GOMP_RTEMS_THREAD_POOLS} set to
4155 @code{"1@@WRK0:3$4@@WRK1"}. Then there are no thread pool restrictions for
4156 scheduler instance @code{IO}. In the scheduler instance @code{WRK0} there is
4157 one thread pool available. Since no priority is specified for this scheduler
4158 instance, the worker thread inherits the priority of the OpenMP primary thread
4159 that created it. In the scheduler instance @code{WRK1} there are three thread
4160 pools available and their worker threads run at priority four.
4165 @c ---------------------------------------------------------------------
4167 @c ---------------------------------------------------------------------
4169 @node Enabling OpenACC
4170 @chapter Enabling OpenACC
4172 To activate the OpenACC extensions for C/C++ and Fortran, the compile-time
4173 flag @option{-fopenacc} must be specified. This enables the OpenACC directive
4174 @samp{#pragma acc} in C/C++ and, in Fortran, the @samp{!$acc} sentinel in free
4175 source form and the @samp{c$acc}, @samp{*$acc} and @samp{!$acc} sentinels in
4176 fixed source form. The flag also arranges for automatic linking of the OpenACC
4177 runtime library (@ref{OpenACC Runtime Library Routines}).
4179 See @uref{https://gcc.gnu.org/wiki/OpenACC} for more information.
4181 A complete description of all OpenACC directives accepted may be found in
4182 the @uref{https://www.openacc.org, OpenACC} Application Programming
4183 Interface manual, version 2.6.
4187 @c ---------------------------------------------------------------------
4188 @c OpenACC Runtime Library Routines
4189 @c ---------------------------------------------------------------------
4191 @node OpenACC Runtime Library Routines
4192 @chapter OpenACC Runtime Library Routines
4194 The runtime routines described here are defined by section 3 of the OpenACC
4195 specifications in version 2.6.
4196 They have C linkage, and do not throw exceptions.
4197 Generally, they are available only for the host, with the exception of
4198 @code{acc_on_device}, which is available for both the host and the
4199 acceleration device.
4202 * acc_get_num_devices:: Get number of devices for the given device
4204 * acc_set_device_type:: Set type of device accelerator to use.
4205 * acc_get_device_type:: Get type of device accelerator to be used.
4206 * acc_set_device_num:: Set device number to use.
4207 * acc_get_device_num:: Get device number to be used.
4208 * acc_get_property:: Get device property.
4209 * acc_async_test:: Tests for completion of a specific asynchronous
4211 * acc_async_test_all:: Tests for completion of all asynchronous
4213 * acc_wait:: Wait for completion of a specific asynchronous
4215 * acc_wait_all:: Waits for completion of all asynchronous
4217 * acc_wait_all_async:: Wait for completion of all asynchronous
4219 * acc_wait_async:: Wait for completion of asynchronous operations.
4220 * acc_init:: Initialize runtime for a specific device type.
4221 * acc_shutdown:: Shuts down the runtime for a specific device
4223 * acc_on_device:: Whether executing on a particular device
4224 * acc_malloc:: Allocate device memory.
4225 * acc_free:: Free device memory.
4226 * acc_copyin:: Allocate device memory and copy host memory to
4228 * acc_present_or_copyin:: If the data is not present on the device,
4229 allocate device memory and copy from host
4231 * acc_create:: Allocate device memory and map it to host
4233 * acc_present_or_create:: If the data is not present on the device,
4234 allocate device memory and map it to host
4236 * acc_copyout:: Copy device memory to host memory.
4237 * acc_delete:: Free device memory.
4238 * acc_update_device:: Update device memory from mapped host memory.
4239 * acc_update_self:: Update host memory from mapped device memory.
4240 * acc_map_data:: Map previously allocated device memory to host
4242 * acc_unmap_data:: Unmap device memory from host memory.
4243 * acc_deviceptr:: Get device pointer associated with specific
4245 * acc_hostptr:: Get host pointer associated with specific
4247 * acc_is_present:: Indicate whether host variable / array is
4249 * acc_memcpy_to_device:: Copy host memory to device memory.
4250 * acc_memcpy_from_device:: Copy device memory to host memory.
4251 * acc_attach:: Let device pointer point to device-pointer target.
4252 * acc_detach:: Let device pointer point to host-pointer target.
4254 API routines for target platforms.
4256 * acc_get_current_cuda_device:: Get CUDA device handle.
4257 * acc_get_current_cuda_context::Get CUDA context handle.
4258 * acc_get_cuda_stream:: Get CUDA stream handle.
4259 * acc_set_cuda_stream:: Set CUDA stream handle.
4261 API routines for the OpenACC Profiling Interface.
4263 * acc_prof_register:: Register callbacks.
4264 * acc_prof_unregister:: Unregister callbacks.
4265 * acc_prof_lookup:: Obtain inquiry functions.
4266 * acc_register_library:: Library registration.
4271 @node acc_get_num_devices
4272 @section @code{acc_get_num_devices} -- Get number of devices for given device type
4274 @item @emph{Description}
4275 This function returns a value indicating the number of devices available
4276 for the device type specified in @var{devicetype}.
4279 @multitable @columnfractions .20 .80
4280 @item @emph{Prototype}: @tab @code{int acc_get_num_devices(acc_device_t devicetype);}
4283 @item @emph{Fortran}:
4284 @multitable @columnfractions .20 .80
4285 @item @emph{Interface}: @tab @code{integer function acc_get_num_devices(devicetype)}
4286 @item @tab @code{integer(kind=acc_device_kind) devicetype}
4289 @item @emph{Reference}:
4290 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4296 @node acc_set_device_type
4297 @section @code{acc_set_device_type} -- Set type of device accelerator to use.
4299 @item @emph{Description}
4300 This function indicates to the runtime library which device type, specified
4301 in @var{devicetype}, to use when executing a parallel or kernels region.
4304 @multitable @columnfractions .20 .80
4305 @item @emph{Prototype}: @tab @code{acc_set_device_type(acc_device_t devicetype);}
4308 @item @emph{Fortran}:
4309 @multitable @columnfractions .20 .80
4310 @item @emph{Interface}: @tab @code{subroutine acc_set_device_type(devicetype)}
4311 @item @tab @code{integer(kind=acc_device_kind) devicetype}
4314 @item @emph{Reference}:
4315 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4321 @node acc_get_device_type
4322 @section @code{acc_get_device_type} -- Get type of device accelerator to be used.
4324 @item @emph{Description}
4325 This function returns what device type will be used when executing a
4326 parallel or kernels region.
4328 This function returns @code{acc_device_none} if
4329 @code{acc_get_device_type} is called from
4330 @code{acc_ev_device_init_start}, @code{acc_ev_device_init_end}
4331 callbacks of the OpenACC Profiling Interface (@ref{OpenACC Profiling
4332 Interface}), that is, if the device is currently being initialized.
4335 @multitable @columnfractions .20 .80
4336 @item @emph{Prototype}: @tab @code{acc_device_t acc_get_device_type(void);}
4339 @item @emph{Fortran}:
4340 @multitable @columnfractions .20 .80
4341 @item @emph{Interface}: @tab @code{function acc_get_device_type(void)}
4342 @item @tab @code{integer(kind=acc_device_kind) acc_get_device_type}
4345 @item @emph{Reference}:
4346 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4352 @node acc_set_device_num
4353 @section @code{acc_set_device_num} -- Set device number to use.
4355 @item @emph{Description}
4356 This function will indicate to the runtime which device number,
4357 specified by @var{devicenum}, associated with the specified device
4358 type @var{devicetype}.
4361 @multitable @columnfractions .20 .80
4362 @item @emph{Prototype}: @tab @code{acc_set_device_num(int devicenum, acc_device_t devicetype);}
4365 @item @emph{Fortran}:
4366 @multitable @columnfractions .20 .80
4367 @item @emph{Interface}: @tab @code{subroutine acc_set_device_num(devicenum, devicetype)}
4368 @item @tab @code{integer devicenum}
4369 @item @tab @code{integer(kind=acc_device_kind) devicetype}
4372 @item @emph{Reference}:
4373 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4379 @node acc_get_device_num
4380 @section @code{acc_get_device_num} -- Get device number to be used.
4382 @item @emph{Description}
4383 This function returns which device number associated with the specified device
4384 type @var{devicetype}, will be used when executing a parallel or kernels
4388 @multitable @columnfractions .20 .80
4389 @item @emph{Prototype}: @tab @code{int acc_get_device_num(acc_device_t devicetype);}
4392 @item @emph{Fortran}:
4393 @multitable @columnfractions .20 .80
4394 @item @emph{Interface}: @tab @code{function acc_get_device_num(devicetype)}
4395 @item @tab @code{integer(kind=acc_device_kind) devicetype}
4396 @item @tab @code{integer acc_get_device_num}
4399 @item @emph{Reference}:
4400 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4406 @node acc_get_property
4407 @section @code{acc_get_property} -- Get device property.
4408 @cindex acc_get_property
4409 @cindex acc_get_property_string
4411 @item @emph{Description}
4412 These routines return the value of the specified @var{property} for the
4413 device being queried according to @var{devicenum} and @var{devicetype}.
4414 Integer-valued and string-valued properties are returned by
4415 @code{acc_get_property} and @code{acc_get_property_string} respectively.
4416 The Fortran @code{acc_get_property_string} subroutine returns the string
4417 retrieved in its fourth argument while the remaining entry points are
4418 functions, which pass the return value as their result.
4420 Note for Fortran, only: the OpenACC technical committee corrected and, hence,
4421 modified the interface introduced in OpenACC 2.6. The kind-value parameter
4422 @code{acc_device_property} has been renamed to @code{acc_device_property_kind}
4423 for consistency and the return type of the @code{acc_get_property} function is
4424 now a @code{c_size_t} integer instead of a @code{acc_device_property} integer.
4425 The parameter @code{acc_device_property} is still provided,
4426 but might be removed in a future version of GCC.
4429 @multitable @columnfractions .20 .80
4430 @item @emph{Prototype}: @tab @code{size_t acc_get_property(int devicenum, acc_device_t devicetype, acc_device_property_t property);}
4431 @item @emph{Prototype}: @tab @code{const char *acc_get_property_string(int devicenum, acc_device_t devicetype, acc_device_property_t property);}
4434 @item @emph{Fortran}:
4435 @multitable @columnfractions .20 .80
4436 @item @emph{Interface}: @tab @code{function acc_get_property(devicenum, devicetype, property)}
4437 @item @emph{Interface}: @tab @code{subroutine acc_get_property_string(devicenum, devicetype, property, string)}
4438 @item @tab @code{use ISO_C_Binding, only: c_size_t}
4439 @item @tab @code{integer devicenum}
4440 @item @tab @code{integer(kind=acc_device_kind) devicetype}
4441 @item @tab @code{integer(kind=acc_device_property_kind) property}
4442 @item @tab @code{integer(kind=c_size_t) acc_get_property}
4443 @item @tab @code{character(*) string}
4446 @item @emph{Reference}:
4447 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4453 @node acc_async_test
4454 @section @code{acc_async_test} -- Test for completion of a specific asynchronous operation.
4456 @item @emph{Description}
4457 This function tests for completion of the asynchronous operation specified
4458 in @var{arg}. In C/C++, a non-zero value is returned to indicate
4459 the specified asynchronous operation has completed while Fortran returns
4460 @code{true}. If the asynchronous operation has not completed, C/C++ returns
4461 zero and Fortran returns @code{false}.
4464 @multitable @columnfractions .20 .80
4465 @item @emph{Prototype}: @tab @code{int acc_async_test(int arg);}
4468 @item @emph{Fortran}:
4469 @multitable @columnfractions .20 .80
4470 @item @emph{Interface}: @tab @code{function acc_async_test(arg)}
4471 @item @tab @code{integer(kind=acc_handle_kind) arg}
4472 @item @tab @code{logical acc_async_test}
4475 @item @emph{Reference}:
4476 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4482 @node acc_async_test_all
4483 @section @code{acc_async_test_all} -- Tests for completion of all asynchronous operations.
4485 @item @emph{Description}
4486 This function tests for completion of all asynchronous operations.
4487 In C/C++, a non-zero value is returned to indicate all asynchronous
4488 operations have completed while Fortran returns @code{true}. If
4489 any asynchronous operation has not completed, C/C++ returns zero and
4490 Fortran returns @code{false}.
4493 @multitable @columnfractions .20 .80
4494 @item @emph{Prototype}: @tab @code{int acc_async_test_all(void);}
4497 @item @emph{Fortran}:
4498 @multitable @columnfractions .20 .80
4499 @item @emph{Interface}: @tab @code{function acc_async_test()}
4500 @item @tab @code{logical acc_get_device_num}
4503 @item @emph{Reference}:
4504 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4511 @section @code{acc_wait} -- Wait for completion of a specific asynchronous operation.
4513 @item @emph{Description}
4514 This function waits for completion of the asynchronous operation
4515 specified in @var{arg}.
4518 @multitable @columnfractions .20 .80
4519 @item @emph{Prototype}: @tab @code{acc_wait(arg);}
4520 @item @emph{Prototype (OpenACC 1.0 compatibility)}: @tab @code{acc_async_wait(arg);}
4523 @item @emph{Fortran}:
4524 @multitable @columnfractions .20 .80
4525 @item @emph{Interface}: @tab @code{subroutine acc_wait(arg)}
4526 @item @tab @code{integer(acc_handle_kind) arg}
4527 @item @emph{Interface (OpenACC 1.0 compatibility)}: @tab @code{subroutine acc_async_wait(arg)}
4528 @item @tab @code{integer(acc_handle_kind) arg}
4531 @item @emph{Reference}:
4532 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4539 @section @code{acc_wait_all} -- Waits for completion of all asynchronous operations.
4541 @item @emph{Description}
4542 This function waits for the completion of all asynchronous operations.
4545 @multitable @columnfractions .20 .80
4546 @item @emph{Prototype}: @tab @code{acc_wait_all(void);}
4547 @item @emph{Prototype (OpenACC 1.0 compatibility)}: @tab @code{acc_async_wait_all(void);}
4550 @item @emph{Fortran}:
4551 @multitable @columnfractions .20 .80
4552 @item @emph{Interface}: @tab @code{subroutine acc_wait_all()}
4553 @item @emph{Interface (OpenACC 1.0 compatibility)}: @tab @code{subroutine acc_async_wait_all()}
4556 @item @emph{Reference}:
4557 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4563 @node acc_wait_all_async
4564 @section @code{acc_wait_all_async} -- Wait for completion of all asynchronous operations.
4566 @item @emph{Description}
4567 This function enqueues a wait operation on the queue @var{async} for any
4568 and all asynchronous operations that have been previously enqueued on
4572 @multitable @columnfractions .20 .80
4573 @item @emph{Prototype}: @tab @code{acc_wait_all_async(int async);}
4576 @item @emph{Fortran}:
4577 @multitable @columnfractions .20 .80
4578 @item @emph{Interface}: @tab @code{subroutine acc_wait_all_async(async)}
4579 @item @tab @code{integer(acc_handle_kind) async}
4582 @item @emph{Reference}:
4583 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4589 @node acc_wait_async
4590 @section @code{acc_wait_async} -- Wait for completion of asynchronous operations.
4592 @item @emph{Description}
4593 This function enqueues a wait operation on queue @var{async} for any and all
4594 asynchronous operations enqueued on queue @var{arg}.
4597 @multitable @columnfractions .20 .80
4598 @item @emph{Prototype}: @tab @code{acc_wait_async(int arg, int async);}
4601 @item @emph{Fortran}:
4602 @multitable @columnfractions .20 .80
4603 @item @emph{Interface}: @tab @code{subroutine acc_wait_async(arg, async)}
4604 @item @tab @code{integer(acc_handle_kind) arg, async}
4607 @item @emph{Reference}:
4608 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4615 @section @code{acc_init} -- Initialize runtime for a specific device type.
4617 @item @emph{Description}
4618 This function initializes the runtime for the device type specified in
4622 @multitable @columnfractions .20 .80
4623 @item @emph{Prototype}: @tab @code{acc_init(acc_device_t devicetype);}
4626 @item @emph{Fortran}:
4627 @multitable @columnfractions .20 .80
4628 @item @emph{Interface}: @tab @code{subroutine acc_init(devicetype)}
4629 @item @tab @code{integer(acc_device_kind) devicetype}
4632 @item @emph{Reference}:
4633 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4640 @section @code{acc_shutdown} -- Shuts down the runtime for a specific device type.
4642 @item @emph{Description}
4643 This function shuts down the runtime for the device type specified in
4647 @multitable @columnfractions .20 .80
4648 @item @emph{Prototype}: @tab @code{acc_shutdown(acc_device_t devicetype);}
4651 @item @emph{Fortran}:
4652 @multitable @columnfractions .20 .80
4653 @item @emph{Interface}: @tab @code{subroutine acc_shutdown(devicetype)}
4654 @item @tab @code{integer(acc_device_kind) devicetype}
4657 @item @emph{Reference}:
4658 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4665 @section @code{acc_on_device} -- Whether executing on a particular device
4667 @item @emph{Description}:
4668 This function returns whether the program is executing on a particular
4669 device specified in @var{devicetype}. In C/C++ a non-zero value is
4670 returned to indicate the device is executing on the specified device type.
4671 In Fortran, @code{true} is returned. If the program is not executing
4672 on the specified device type C/C++ returns zero, while Fortran
4673 returns @code{false}.
4676 @multitable @columnfractions .20 .80
4677 @item @emph{Prototype}: @tab @code{acc_on_device(acc_device_t devicetype);}
4680 @item @emph{Fortran}:
4681 @multitable @columnfractions .20 .80
4682 @item @emph{Interface}: @tab @code{function acc_on_device(devicetype)}
4683 @item @tab @code{integer(acc_device_kind) devicetype}
4684 @item @tab @code{logical acc_on_device}
4688 @item @emph{Reference}:
4689 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4696 @section @code{acc_malloc} -- Allocate device memory.
4698 @item @emph{Description}
4699 This function allocates @var{len} bytes of device memory. It returns
4700 the device address of the allocated memory.
4703 @multitable @columnfractions .20 .80
4704 @item @emph{Prototype}: @tab @code{d_void* acc_malloc(size_t len);}
4707 @item @emph{Reference}:
4708 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4715 @section @code{acc_free} -- Free device memory.
4717 @item @emph{Description}
4718 Free previously allocated device memory at the device address @code{a}.
4721 @multitable @columnfractions .20 .80
4722 @item @emph{Prototype}: @tab @code{acc_free(d_void *a);}
4725 @item @emph{Reference}:
4726 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4733 @section @code{acc_copyin} -- Allocate device memory and copy host memory to it.
4735 @item @emph{Description}
4736 In C/C++, this function allocates @var{len} bytes of device memory
4737 and maps it to the specified host address in @var{a}. The device
4738 address of the newly allocated device memory is returned.
4740 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
4741 a contiguous array section. The second form @var{a} specifies a
4742 variable or array element and @var{len} specifies the length in bytes.
4745 @multitable @columnfractions .20 .80
4746 @item @emph{Prototype}: @tab @code{void *acc_copyin(h_void *a, size_t len);}
4747 @item @emph{Prototype}: @tab @code{void *acc_copyin_async(h_void *a, size_t len, int async);}
4750 @item @emph{Fortran}:
4751 @multitable @columnfractions .20 .80
4752 @item @emph{Interface}: @tab @code{subroutine acc_copyin(a)}
4753 @item @tab @code{type, dimension(:[,:]...) :: a}
4754 @item @emph{Interface}: @tab @code{subroutine acc_copyin(a, len)}
4755 @item @tab @code{type, dimension(:[,:]...) :: a}
4756 @item @tab @code{integer len}
4757 @item @emph{Interface}: @tab @code{subroutine acc_copyin_async(a, async)}
4758 @item @tab @code{type, dimension(:[,:]...) :: a}
4759 @item @tab @code{integer(acc_handle_kind) :: async}
4760 @item @emph{Interface}: @tab @code{subroutine acc_copyin_async(a, len, async)}
4761 @item @tab @code{type, dimension(:[,:]...) :: a}
4762 @item @tab @code{integer len}
4763 @item @tab @code{integer(acc_handle_kind) :: async}
4766 @item @emph{Reference}:
4767 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4773 @node acc_present_or_copyin
4774 @section @code{acc_present_or_copyin} -- If the data is not present on the device, allocate device memory and copy from host memory.
4776 @item @emph{Description}
4777 This function tests if the host data specified by @var{a} and of length
4778 @var{len} is present or not. If it is not present, device memory
4779 is allocated and the host memory copied. The device address of
4780 the newly allocated device memory is returned.
4782 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
4783 a contiguous array section. The second form @var{a} specifies a variable or
4784 array element and @var{len} specifies the length in bytes.
4786 Note that @code{acc_present_or_copyin} and @code{acc_pcopyin} exist for
4787 backward compatibility with OpenACC 2.0; use @ref{acc_copyin} instead.
4790 @multitable @columnfractions .20 .80
4791 @item @emph{Prototype}: @tab @code{void *acc_present_or_copyin(h_void *a, size_t len);}
4792 @item @emph{Prototype}: @tab @code{void *acc_pcopyin(h_void *a, size_t len);}
4795 @item @emph{Fortran}:
4796 @multitable @columnfractions .20 .80
4797 @item @emph{Interface}: @tab @code{subroutine acc_present_or_copyin(a)}
4798 @item @tab @code{type, dimension(:[,:]...) :: a}
4799 @item @emph{Interface}: @tab @code{subroutine acc_present_or_copyin(a, len)}
4800 @item @tab @code{type, dimension(:[,:]...) :: a}
4801 @item @tab @code{integer len}
4802 @item @emph{Interface}: @tab @code{subroutine acc_pcopyin(a)}
4803 @item @tab @code{type, dimension(:[,:]...) :: a}
4804 @item @emph{Interface}: @tab @code{subroutine acc_pcopyin(a, len)}
4805 @item @tab @code{type, dimension(:[,:]...) :: a}
4806 @item @tab @code{integer len}
4809 @item @emph{Reference}:
4810 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4817 @section @code{acc_create} -- Allocate device memory and map it to host memory.
4819 @item @emph{Description}
4820 This function allocates device memory and maps it to host memory specified
4821 by the host address @var{a} with a length of @var{len} bytes. In C/C++,
4822 the function returns the device address of the allocated device memory.
4824 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
4825 a contiguous array section. The second form @var{a} specifies a variable or
4826 array element and @var{len} specifies the length in bytes.
4829 @multitable @columnfractions .20 .80
4830 @item @emph{Prototype}: @tab @code{void *acc_create(h_void *a, size_t len);}
4831 @item @emph{Prototype}: @tab @code{void *acc_create_async(h_void *a, size_t len, int async);}
4834 @item @emph{Fortran}:
4835 @multitable @columnfractions .20 .80
4836 @item @emph{Interface}: @tab @code{subroutine acc_create(a)}
4837 @item @tab @code{type, dimension(:[,:]...) :: a}
4838 @item @emph{Interface}: @tab @code{subroutine acc_create(a, len)}
4839 @item @tab @code{type, dimension(:[,:]...) :: a}
4840 @item @tab @code{integer len}
4841 @item @emph{Interface}: @tab @code{subroutine acc_create_async(a, async)}
4842 @item @tab @code{type, dimension(:[,:]...) :: a}
4843 @item @tab @code{integer(acc_handle_kind) :: async}
4844 @item @emph{Interface}: @tab @code{subroutine acc_create_async(a, len, async)}
4845 @item @tab @code{type, dimension(:[,:]...) :: a}
4846 @item @tab @code{integer len}
4847 @item @tab @code{integer(acc_handle_kind) :: async}
4850 @item @emph{Reference}:
4851 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4857 @node acc_present_or_create
4858 @section @code{acc_present_or_create} -- If the data is not present on the device, allocate device memory and map it to host memory.
4860 @item @emph{Description}
4861 This function tests if the host data specified by @var{a} and of length
4862 @var{len} is present or not. If it is not present, device memory
4863 is allocated and mapped to host memory. In C/C++, the device address
4864 of the newly allocated device memory is returned.
4866 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
4867 a contiguous array section. The second form @var{a} specifies a variable or
4868 array element and @var{len} specifies the length in bytes.
4870 Note that @code{acc_present_or_create} and @code{acc_pcreate} exist for
4871 backward compatibility with OpenACC 2.0; use @ref{acc_create} instead.
4874 @multitable @columnfractions .20 .80
4875 @item @emph{Prototype}: @tab @code{void *acc_present_or_create(h_void *a, size_t len)}
4876 @item @emph{Prototype}: @tab @code{void *acc_pcreate(h_void *a, size_t len)}
4879 @item @emph{Fortran}:
4880 @multitable @columnfractions .20 .80
4881 @item @emph{Interface}: @tab @code{subroutine acc_present_or_create(a)}
4882 @item @tab @code{type, dimension(:[,:]...) :: a}
4883 @item @emph{Interface}: @tab @code{subroutine acc_present_or_create(a, len)}
4884 @item @tab @code{type, dimension(:[,:]...) :: a}
4885 @item @tab @code{integer len}
4886 @item @emph{Interface}: @tab @code{subroutine acc_pcreate(a)}
4887 @item @tab @code{type, dimension(:[,:]...) :: a}
4888 @item @emph{Interface}: @tab @code{subroutine acc_pcreate(a, len)}
4889 @item @tab @code{type, dimension(:[,:]...) :: a}
4890 @item @tab @code{integer len}
4893 @item @emph{Reference}:
4894 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4901 @section @code{acc_copyout} -- Copy device memory to host memory.
4903 @item @emph{Description}
4904 This function copies mapped device memory to host memory which is specified
4905 by host address @var{a} for a length @var{len} bytes in C/C++.
4907 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
4908 a contiguous array section. The second form @var{a} specifies a variable or
4909 array element and @var{len} specifies the length in bytes.
4912 @multitable @columnfractions .20 .80
4913 @item @emph{Prototype}: @tab @code{acc_copyout(h_void *a, size_t len);}
4914 @item @emph{Prototype}: @tab @code{acc_copyout_async(h_void *a, size_t len, int async);}
4915 @item @emph{Prototype}: @tab @code{acc_copyout_finalize(h_void *a, size_t len);}
4916 @item @emph{Prototype}: @tab @code{acc_copyout_finalize_async(h_void *a, size_t len, int async);}
4919 @item @emph{Fortran}:
4920 @multitable @columnfractions .20 .80
4921 @item @emph{Interface}: @tab @code{subroutine acc_copyout(a)}
4922 @item @tab @code{type, dimension(:[,:]...) :: a}
4923 @item @emph{Interface}: @tab @code{subroutine acc_copyout(a, len)}
4924 @item @tab @code{type, dimension(:[,:]...) :: a}
4925 @item @tab @code{integer len}
4926 @item @emph{Interface}: @tab @code{subroutine acc_copyout_async(a, async)}
4927 @item @tab @code{type, dimension(:[,:]...) :: a}
4928 @item @tab @code{integer(acc_handle_kind) :: async}
4929 @item @emph{Interface}: @tab @code{subroutine acc_copyout_async(a, len, async)}
4930 @item @tab @code{type, dimension(:[,:]...) :: a}
4931 @item @tab @code{integer len}
4932 @item @tab @code{integer(acc_handle_kind) :: async}
4933 @item @emph{Interface}: @tab @code{subroutine acc_copyout_finalize(a)}
4934 @item @tab @code{type, dimension(:[,:]...) :: a}
4935 @item @emph{Interface}: @tab @code{subroutine acc_copyout_finalize(a, len)}
4936 @item @tab @code{type, dimension(:[,:]...) :: a}
4937 @item @tab @code{integer len}
4938 @item @emph{Interface}: @tab @code{subroutine acc_copyout_finalize_async(a, async)}
4939 @item @tab @code{type, dimension(:[,:]...) :: a}
4940 @item @tab @code{integer(acc_handle_kind) :: async}
4941 @item @emph{Interface}: @tab @code{subroutine acc_copyout_finalize_async(a, len, async)}
4942 @item @tab @code{type, dimension(:[,:]...) :: a}
4943 @item @tab @code{integer len}
4944 @item @tab @code{integer(acc_handle_kind) :: async}
4947 @item @emph{Reference}:
4948 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
4955 @section @code{acc_delete} -- Free device memory.
4957 @item @emph{Description}
4958 This function frees previously allocated device memory specified by
4959 the device address @var{a} and the length of @var{len} bytes.
4961 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
4962 a contiguous array section. The second form @var{a} specifies a variable or
4963 array element and @var{len} specifies the length in bytes.
4966 @multitable @columnfractions .20 .80
4967 @item @emph{Prototype}: @tab @code{acc_delete(h_void *a, size_t len);}
4968 @item @emph{Prototype}: @tab @code{acc_delete_async(h_void *a, size_t len, int async);}
4969 @item @emph{Prototype}: @tab @code{acc_delete_finalize(h_void *a, size_t len);}
4970 @item @emph{Prototype}: @tab @code{acc_delete_finalize_async(h_void *a, size_t len, int async);}
4973 @item @emph{Fortran}:
4974 @multitable @columnfractions .20 .80
4975 @item @emph{Interface}: @tab @code{subroutine acc_delete(a)}
4976 @item @tab @code{type, dimension(:[,:]...) :: a}
4977 @item @emph{Interface}: @tab @code{subroutine acc_delete(a, len)}
4978 @item @tab @code{type, dimension(:[,:]...) :: a}
4979 @item @tab @code{integer len}
4980 @item @emph{Interface}: @tab @code{subroutine acc_delete_async(a, async)}
4981 @item @tab @code{type, dimension(:[,:]...) :: a}
4982 @item @tab @code{integer(acc_handle_kind) :: async}
4983 @item @emph{Interface}: @tab @code{subroutine acc_delete_async(a, len, async)}
4984 @item @tab @code{type, dimension(:[,:]...) :: a}
4985 @item @tab @code{integer len}
4986 @item @tab @code{integer(acc_handle_kind) :: async}
4987 @item @emph{Interface}: @tab @code{subroutine acc_delete_finalize(a)}
4988 @item @tab @code{type, dimension(:[,:]...) :: a}
4989 @item @emph{Interface}: @tab @code{subroutine acc_delete_finalize(a, len)}
4990 @item @tab @code{type, dimension(:[,:]...) :: a}
4991 @item @tab @code{integer len}
4992 @item @emph{Interface}: @tab @code{subroutine acc_delete_async_finalize(a, async)}
4993 @item @tab @code{type, dimension(:[,:]...) :: a}
4994 @item @tab @code{integer(acc_handle_kind) :: async}
4995 @item @emph{Interface}: @tab @code{subroutine acc_delete_async_finalize(a, len, async)}
4996 @item @tab @code{type, dimension(:[,:]...) :: a}
4997 @item @tab @code{integer len}
4998 @item @tab @code{integer(acc_handle_kind) :: async}
5001 @item @emph{Reference}:
5002 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5008 @node acc_update_device
5009 @section @code{acc_update_device} -- Update device memory from mapped host memory.
5011 @item @emph{Description}
5012 This function updates the device copy from the previously mapped host memory.
5013 The host memory is specified with the host address @var{a} and a length of
5016 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
5017 a contiguous array section. The second form @var{a} specifies a variable or
5018 array element and @var{len} specifies the length in bytes.
5021 @multitable @columnfractions .20 .80
5022 @item @emph{Prototype}: @tab @code{acc_update_device(h_void *a, size_t len);}
5023 @item @emph{Prototype}: @tab @code{acc_update_device(h_void *a, size_t len, async);}
5026 @item @emph{Fortran}:
5027 @multitable @columnfractions .20 .80
5028 @item @emph{Interface}: @tab @code{subroutine acc_update_device(a)}
5029 @item @tab @code{type, dimension(:[,:]...) :: a}
5030 @item @emph{Interface}: @tab @code{subroutine acc_update_device(a, len)}
5031 @item @tab @code{type, dimension(:[,:]...) :: a}
5032 @item @tab @code{integer len}
5033 @item @emph{Interface}: @tab @code{subroutine acc_update_device_async(a, async)}
5034 @item @tab @code{type, dimension(:[,:]...) :: a}
5035 @item @tab @code{integer(acc_handle_kind) :: async}
5036 @item @emph{Interface}: @tab @code{subroutine acc_update_device_async(a, len, async)}
5037 @item @tab @code{type, dimension(:[,:]...) :: a}
5038 @item @tab @code{integer len}
5039 @item @tab @code{integer(acc_handle_kind) :: async}
5042 @item @emph{Reference}:
5043 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5049 @node acc_update_self
5050 @section @code{acc_update_self} -- Update host memory from mapped device memory.
5052 @item @emph{Description}
5053 This function updates the host copy from the previously mapped device memory.
5054 The host memory is specified with the host address @var{a} and a length of
5057 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
5058 a contiguous array section. The second form @var{a} specifies a variable or
5059 array element and @var{len} specifies the length in bytes.
5062 @multitable @columnfractions .20 .80
5063 @item @emph{Prototype}: @tab @code{acc_update_self(h_void *a, size_t len);}
5064 @item @emph{Prototype}: @tab @code{acc_update_self_async(h_void *a, size_t len, int async);}
5067 @item @emph{Fortran}:
5068 @multitable @columnfractions .20 .80
5069 @item @emph{Interface}: @tab @code{subroutine acc_update_self(a)}
5070 @item @tab @code{type, dimension(:[,:]...) :: a}
5071 @item @emph{Interface}: @tab @code{subroutine acc_update_self(a, len)}
5072 @item @tab @code{type, dimension(:[,:]...) :: a}
5073 @item @tab @code{integer len}
5074 @item @emph{Interface}: @tab @code{subroutine acc_update_self_async(a, async)}
5075 @item @tab @code{type, dimension(:[,:]...) :: a}
5076 @item @tab @code{integer(acc_handle_kind) :: async}
5077 @item @emph{Interface}: @tab @code{subroutine acc_update_self_async(a, len, async)}
5078 @item @tab @code{type, dimension(:[,:]...) :: a}
5079 @item @tab @code{integer len}
5080 @item @tab @code{integer(acc_handle_kind) :: async}
5083 @item @emph{Reference}:
5084 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5091 @section @code{acc_map_data} -- Map previously allocated device memory to host memory.
5093 @item @emph{Description}
5094 This function maps previously allocated device and host memory. The device
5095 memory is specified with the device address @var{d}. The host memory is
5096 specified with the host address @var{h} and a length of @var{len}.
5099 @multitable @columnfractions .20 .80
5100 @item @emph{Prototype}: @tab @code{acc_map_data(h_void *h, d_void *d, size_t len);}
5103 @item @emph{Reference}:
5104 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5110 @node acc_unmap_data
5111 @section @code{acc_unmap_data} -- Unmap device memory from host memory.
5113 @item @emph{Description}
5114 This function unmaps previously mapped device and host memory. The latter
5115 specified by @var{h}.
5118 @multitable @columnfractions .20 .80
5119 @item @emph{Prototype}: @tab @code{acc_unmap_data(h_void *h);}
5122 @item @emph{Reference}:
5123 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5130 @section @code{acc_deviceptr} -- Get device pointer associated with specific host address.
5132 @item @emph{Description}
5133 This function returns the device address that has been mapped to the
5134 host address specified by @var{h}.
5137 @multitable @columnfractions .20 .80
5138 @item @emph{Prototype}: @tab @code{void *acc_deviceptr(h_void *h);}
5141 @item @emph{Reference}:
5142 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5149 @section @code{acc_hostptr} -- Get host pointer associated with specific device address.
5151 @item @emph{Description}
5152 This function returns the host address that has been mapped to the
5153 device address specified by @var{d}.
5156 @multitable @columnfractions .20 .80
5157 @item @emph{Prototype}: @tab @code{void *acc_hostptr(d_void *d);}
5160 @item @emph{Reference}:
5161 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5167 @node acc_is_present
5168 @section @code{acc_is_present} -- Indicate whether host variable / array is present on device.
5170 @item @emph{Description}
5171 This function indicates whether the specified host address in @var{a} and a
5172 length of @var{len} bytes is present on the device. In C/C++, a non-zero
5173 value is returned to indicate the presence of the mapped memory on the
5174 device. A zero is returned to indicate the memory is not mapped on the
5177 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
5178 a contiguous array section. The second form @var{a} specifies a variable or
5179 array element and @var{len} specifies the length in bytes. If the host
5180 memory is mapped to device memory, then a @code{true} is returned. Otherwise,
5181 a @code{false} is return to indicate the mapped memory is not present.
5184 @multitable @columnfractions .20 .80
5185 @item @emph{Prototype}: @tab @code{int acc_is_present(h_void *a, size_t len);}
5188 @item @emph{Fortran}:
5189 @multitable @columnfractions .20 .80
5190 @item @emph{Interface}: @tab @code{function acc_is_present(a)}
5191 @item @tab @code{type, dimension(:[,:]...) :: a}
5192 @item @tab @code{logical acc_is_present}
5193 @item @emph{Interface}: @tab @code{function acc_is_present(a, len)}
5194 @item @tab @code{type, dimension(:[,:]...) :: a}
5195 @item @tab @code{integer len}
5196 @item @tab @code{logical acc_is_present}
5199 @item @emph{Reference}:
5200 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5206 @node acc_memcpy_to_device
5207 @section @code{acc_memcpy_to_device} -- Copy host memory to device memory.
5209 @item @emph{Description}
5210 This function copies host memory specified by host address of @var{src} to
5211 device memory specified by the device address @var{dest} for a length of
5215 @multitable @columnfractions .20 .80
5216 @item @emph{Prototype}: @tab @code{acc_memcpy_to_device(d_void *dest, h_void *src, size_t bytes);}
5219 @item @emph{Reference}:
5220 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5226 @node acc_memcpy_from_device
5227 @section @code{acc_memcpy_from_device} -- Copy device memory to host memory.
5229 @item @emph{Description}
5230 This function copies host memory specified by host address of @var{src} from
5231 device memory specified by the device address @var{dest} for a length of
5235 @multitable @columnfractions .20 .80
5236 @item @emph{Prototype}: @tab @code{acc_memcpy_from_device(d_void *dest, h_void *src, size_t bytes);}
5239 @item @emph{Reference}:
5240 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5247 @section @code{acc_attach} -- Let device pointer point to device-pointer target.
5249 @item @emph{Description}
5250 This function updates a pointer on the device from pointing to a host-pointer
5251 address to pointing to the corresponding device data.
5254 @multitable @columnfractions .20 .80
5255 @item @emph{Prototype}: @tab @code{acc_attach(h_void **ptr);}
5256 @item @emph{Prototype}: @tab @code{acc_attach_async(h_void **ptr, int async);}
5259 @item @emph{Reference}:
5260 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5267 @section @code{acc_detach} -- Let device pointer point to host-pointer target.
5269 @item @emph{Description}
5270 This function updates a pointer on the device from pointing to a device-pointer
5271 address to pointing to the corresponding host data.
5274 @multitable @columnfractions .20 .80
5275 @item @emph{Prototype}: @tab @code{acc_detach(h_void **ptr);}
5276 @item @emph{Prototype}: @tab @code{acc_detach_async(h_void **ptr, int async);}
5277 @item @emph{Prototype}: @tab @code{acc_detach_finalize(h_void **ptr);}
5278 @item @emph{Prototype}: @tab @code{acc_detach_finalize_async(h_void **ptr, int async);}
5281 @item @emph{Reference}:
5282 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5288 @node acc_get_current_cuda_device
5289 @section @code{acc_get_current_cuda_device} -- Get CUDA device handle.
5291 @item @emph{Description}
5292 This function returns the CUDA device handle. This handle is the same
5293 as used by the CUDA Runtime or Driver API's.
5296 @multitable @columnfractions .20 .80
5297 @item @emph{Prototype}: @tab @code{void *acc_get_current_cuda_device(void);}
5300 @item @emph{Reference}:
5301 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5307 @node acc_get_current_cuda_context
5308 @section @code{acc_get_current_cuda_context} -- Get CUDA context handle.
5310 @item @emph{Description}
5311 This function returns the CUDA context handle. This handle is the same
5312 as used by the CUDA Runtime or Driver API's.
5315 @multitable @columnfractions .20 .80
5316 @item @emph{Prototype}: @tab @code{void *acc_get_current_cuda_context(void);}
5319 @item @emph{Reference}:
5320 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5326 @node acc_get_cuda_stream
5327 @section @code{acc_get_cuda_stream} -- Get CUDA stream handle.
5329 @item @emph{Description}
5330 This function returns the CUDA stream handle for the queue @var{async}.
5331 This handle is the same as used by the CUDA Runtime or Driver API's.
5334 @multitable @columnfractions .20 .80
5335 @item @emph{Prototype}: @tab @code{void *acc_get_cuda_stream(int async);}
5338 @item @emph{Reference}:
5339 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5345 @node acc_set_cuda_stream
5346 @section @code{acc_set_cuda_stream} -- Set CUDA stream handle.
5348 @item @emph{Description}
5349 This function associates the stream handle specified by @var{stream} with
5350 the queue @var{async}.
5352 This cannot be used to change the stream handle associated with
5353 @code{acc_async_sync}.
5355 The return value is not specified.
5358 @multitable @columnfractions .20 .80
5359 @item @emph{Prototype}: @tab @code{int acc_set_cuda_stream(int async, void *stream);}
5362 @item @emph{Reference}:
5363 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5369 @node acc_prof_register
5370 @section @code{acc_prof_register} -- Register callbacks.
5372 @item @emph{Description}:
5373 This function registers callbacks.
5376 @multitable @columnfractions .20 .80
5377 @item @emph{Prototype}: @tab @code{void acc_prof_register (acc_event_t, acc_prof_callback, acc_register_t);}
5380 @item @emph{See also}:
5381 @ref{OpenACC Profiling Interface}
5383 @item @emph{Reference}:
5384 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5390 @node acc_prof_unregister
5391 @section @code{acc_prof_unregister} -- Unregister callbacks.
5393 @item @emph{Description}:
5394 This function unregisters callbacks.
5397 @multitable @columnfractions .20 .80
5398 @item @emph{Prototype}: @tab @code{void acc_prof_unregister (acc_event_t, acc_prof_callback, acc_register_t);}
5401 @item @emph{See also}:
5402 @ref{OpenACC Profiling Interface}
5404 @item @emph{Reference}:
5405 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5411 @node acc_prof_lookup
5412 @section @code{acc_prof_lookup} -- Obtain inquiry functions.
5414 @item @emph{Description}:
5415 Function to obtain inquiry functions.
5418 @multitable @columnfractions .20 .80
5419 @item @emph{Prototype}: @tab @code{acc_query_fn acc_prof_lookup (const char *);}
5422 @item @emph{See also}:
5423 @ref{OpenACC Profiling Interface}
5425 @item @emph{Reference}:
5426 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5432 @node acc_register_library
5433 @section @code{acc_register_library} -- Library registration.
5435 @item @emph{Description}:
5436 Function for library registration.
5439 @multitable @columnfractions .20 .80
5440 @item @emph{Prototype}: @tab @code{void acc_register_library (acc_prof_reg, acc_prof_reg, acc_prof_lookup_func);}
5443 @item @emph{See also}:
5444 @ref{OpenACC Profiling Interface}, @ref{ACC_PROFLIB}
5446 @item @emph{Reference}:
5447 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5453 @c ---------------------------------------------------------------------
5454 @c OpenACC Environment Variables
5455 @c ---------------------------------------------------------------------
5457 @node OpenACC Environment Variables
5458 @chapter OpenACC Environment Variables
5460 The variables @env{ACC_DEVICE_TYPE} and @env{ACC_DEVICE_NUM}
5461 are defined by section 4 of the OpenACC specification in version 2.0.
5462 The variable @env{ACC_PROFLIB}
5463 is defined by section 4 of the OpenACC specification in version 2.6.
5473 @node ACC_DEVICE_TYPE
5474 @section @code{ACC_DEVICE_TYPE}
5476 @item @emph{Description}:
5477 Control the default device type to use when executing compute regions.
5478 If unset, the code can be run on any device type, favoring a non-host
5481 Supported values in GCC (if compiled in) are
5487 @item @emph{Reference}:
5488 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5494 @node ACC_DEVICE_NUM
5495 @section @code{ACC_DEVICE_NUM}
5497 @item @emph{Description}:
5498 Control which device, identified by device number, is the default device.
5499 The value must be a nonnegative integer less than the number of devices.
5500 If unset, device number zero is used.
5501 @item @emph{Reference}:
5502 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5509 @section @code{ACC_PROFLIB}
5511 @item @emph{Description}:
5512 Semicolon-separated list of dynamic libraries that are loaded as profiling
5513 libraries. Each library must provide at least the @code{acc_register_library}
5514 routine. Each library file is found as described by the documentation of
5515 @code{dlopen} of your operating system.
5516 @item @emph{See also}:
5517 @ref{acc_register_library}, @ref{OpenACC Profiling Interface}
5519 @item @emph{Reference}:
5520 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
5526 @c ---------------------------------------------------------------------
5527 @c CUDA Streams Usage
5528 @c ---------------------------------------------------------------------
5530 @node CUDA Streams Usage
5531 @chapter CUDA Streams Usage
5533 This applies to the @code{nvptx} plugin only.
5535 The library provides elements that perform asynchronous movement of
5536 data and asynchronous operation of computing constructs. This
5537 asynchronous functionality is implemented by making use of CUDA
5538 streams@footnote{See "Stream Management" in "CUDA Driver API",
5539 TRM-06703-001, Version 5.5, for additional information}.
5541 The primary means by that the asynchronous functionality is accessed
5542 is through the use of those OpenACC directives which make use of the
5543 @code{async} and @code{wait} clauses. When the @code{async} clause is
5544 first used with a directive, it creates a CUDA stream. If an
5545 @code{async-argument} is used with the @code{async} clause, then the
5546 stream is associated with the specified @code{async-argument}.
5548 Following the creation of an association between a CUDA stream and the
5549 @code{async-argument} of an @code{async} clause, both the @code{wait}
5550 clause and the @code{wait} directive can be used. When either the
5551 clause or directive is used after stream creation, it creates a
5552 rendezvous point whereby execution waits until all operations
5553 associated with the @code{async-argument}, that is, stream, have
5556 Normally, the management of the streams that are created as a result of
5557 using the @code{async} clause, is done without any intervention by the
5558 caller. This implies the association between the @code{async-argument}
5559 and the CUDA stream is maintained for the lifetime of the program.
5560 However, this association can be changed through the use of the library
5561 function @code{acc_set_cuda_stream}. When the function
5562 @code{acc_set_cuda_stream} is called, the CUDA stream that was
5563 originally associated with the @code{async} clause is destroyed.
5564 Caution should be taken when changing the association as subsequent
5565 references to the @code{async-argument} refer to a different
5570 @c ---------------------------------------------------------------------
5571 @c OpenACC Library Interoperability
5572 @c ---------------------------------------------------------------------
5574 @node OpenACC Library Interoperability
5575 @chapter OpenACC Library Interoperability
5577 @section Introduction
5579 The OpenACC library uses the CUDA Driver API, and may interact with
5580 programs that use the Runtime library directly, or another library
5581 based on the Runtime library, e.g., CUBLAS@footnote{See section 2.26,
5582 "Interactions with the CUDA Driver API" in
5583 "CUDA Runtime API", Version 5.5, and section 2.27, "VDPAU
5584 Interoperability", in "CUDA Driver API", TRM-06703-001, Version 5.5,
5585 for additional information on library interoperability.}.
5586 This chapter describes the use cases and what changes are
5587 required in order to use both the OpenACC library and the CUBLAS and Runtime
5588 libraries within a program.
5590 @section First invocation: NVIDIA CUBLAS library API
5592 In this first use case (see below), a function in the CUBLAS library is called
5593 prior to any of the functions in the OpenACC library. More specifically, the
5594 function @code{cublasCreate()}.
5596 When invoked, the function initializes the library and allocates the
5597 hardware resources on the host and the device on behalf of the caller. Once
5598 the initialization and allocation has completed, a handle is returned to the
5599 caller. The OpenACC library also requires initialization and allocation of
5600 hardware resources. Since the CUBLAS library has already allocated the
5601 hardware resources for the device, all that is left to do is to initialize
5602 the OpenACC library and acquire the hardware resources on the host.
5604 Prior to calling the OpenACC function that initializes the library and
5605 allocate the host hardware resources, you need to acquire the device number
5606 that was allocated during the call to @code{cublasCreate()}. The invoking of the
5607 runtime library function @code{cudaGetDevice()} accomplishes this. Once
5608 acquired, the device number is passed along with the device type as
5609 parameters to the OpenACC library function @code{acc_set_device_num()}.
5611 Once the call to @code{acc_set_device_num()} has completed, the OpenACC
5612 library uses the context that was created during the call to
5613 @code{cublasCreate()}. In other words, both libraries share the
5617 /* Create the handle */
5618 s = cublasCreate(&h);
5619 if (s != CUBLAS_STATUS_SUCCESS)
5621 fprintf(stderr, "cublasCreate failed %d\n", s);
5625 /* Get the device number */
5626 e = cudaGetDevice(&dev);
5627 if (e != cudaSuccess)
5629 fprintf(stderr, "cudaGetDevice failed %d\n", e);
5633 /* Initialize OpenACC library and use device 'dev' */
5634 acc_set_device_num(dev, acc_device_nvidia);
5639 @section First invocation: OpenACC library API
5641 In this second use case (see below), a function in the OpenACC library is
5642 called prior to any of the functions in the CUBLAS library. More specifically,
5643 the function @code{acc_set_device_num()}.
5645 In the use case presented here, the function @code{acc_set_device_num()}
5646 is used to both initialize the OpenACC library and allocate the hardware
5647 resources on the host and the device. In the call to the function, the
5648 call parameters specify which device to use and what device
5649 type to use, i.e., @code{acc_device_nvidia}. It should be noted that this
5650 is but one method to initialize the OpenACC library and allocate the
5651 appropriate hardware resources. Other methods are available through the
5652 use of environment variables and these is discussed in the next section.
5654 Once the call to @code{acc_set_device_num()} has completed, other OpenACC
5655 functions can be called as seen with multiple calls being made to
5656 @code{acc_copyin()}. In addition, calls can be made to functions in the
5657 CUBLAS library. In the use case a call to @code{cublasCreate()} is made
5658 subsequent to the calls to @code{acc_copyin()}.
5659 As seen in the previous use case, a call to @code{cublasCreate()}
5660 initializes the CUBLAS library and allocates the hardware resources on the
5661 host and the device. However, since the device has already been allocated,
5662 @code{cublasCreate()} only initializes the CUBLAS library and allocates
5663 the appropriate hardware resources on the host. The context that was created
5664 as part of the OpenACC initialization is shared with the CUBLAS library,
5665 similarly to the first use case.
5670 acc_set_device_num(dev, acc_device_nvidia);
5672 /* Copy the first set to the device */
5673 d_X = acc_copyin(&h_X[0], N * sizeof (float));
5676 fprintf(stderr, "copyin error h_X\n");
5680 /* Copy the second set to the device */
5681 d_Y = acc_copyin(&h_Y1[0], N * sizeof (float));
5684 fprintf(stderr, "copyin error h_Y1\n");
5688 /* Create the handle */
5689 s = cublasCreate(&h);
5690 if (s != CUBLAS_STATUS_SUCCESS)
5692 fprintf(stderr, "cublasCreate failed %d\n", s);
5696 /* Perform saxpy using CUBLAS library function */
5697 s = cublasSaxpy(h, N, &alpha, d_X, 1, d_Y, 1);
5698 if (s != CUBLAS_STATUS_SUCCESS)
5700 fprintf(stderr, "cublasSaxpy failed %d\n", s);
5704 /* Copy the results from the device */
5705 acc_memcpy_from_device(&h_Y1[0], d_Y, N * sizeof (float));
5710 @section OpenACC library and environment variables
5712 There are two environment variables associated with the OpenACC library
5713 that may be used to control the device type and device number:
5714 @env{ACC_DEVICE_TYPE} and @env{ACC_DEVICE_NUM}, respectively. These two
5715 environment variables can be used as an alternative to calling
5716 @code{acc_set_device_num()}. As seen in the second use case, the device
5717 type and device number were specified using @code{acc_set_device_num()}.
5718 If however, the aforementioned environment variables were set, then the
5719 call to @code{acc_set_device_num()} would not be required.
5722 The use of the environment variables is only relevant when an OpenACC function
5723 is called prior to a call to @code{cudaCreate()}. If @code{cudaCreate()}
5724 is called prior to a call to an OpenACC function, then you must call
5725 @code{acc_set_device_num()}@footnote{More complete information
5726 about @env{ACC_DEVICE_TYPE} and @env{ACC_DEVICE_NUM} can be found in
5727 sections 4.1 and 4.2 of the @uref{https://www.openacc.org, OpenACC}
5728 Application Programming Interface”, Version 2.6.}
5732 @c ---------------------------------------------------------------------
5733 @c OpenACC Profiling Interface
5734 @c ---------------------------------------------------------------------
5736 @node OpenACC Profiling Interface
5737 @chapter OpenACC Profiling Interface
5739 @section Implementation Status and Implementation-Defined Behavior
5741 We're implementing the OpenACC Profiling Interface as defined by the
5742 OpenACC 2.6 specification. We're clarifying some aspects here as
5743 @emph{implementation-defined behavior}, while they're still under
5744 discussion within the OpenACC Technical Committee.
5746 This implementation is tuned to keep the performance impact as low as
5747 possible for the (very common) case that the Profiling Interface is
5748 not enabled. This is relevant, as the Profiling Interface affects all
5749 the @emph{hot} code paths (in the target code, not in the offloaded
5750 code). Users of the OpenACC Profiling Interface can be expected to
5751 understand that performance is impacted to some degree once the
5752 Profiling Interface is enabled: for example, because of the
5753 @emph{runtime} (libgomp) calling into a third-party @emph{library} for
5754 every event that has been registered.
5756 We're not yet accounting for the fact that @cite{OpenACC events may
5757 occur during event processing}.
5758 We just handle one case specially, as required by CUDA 9.0
5759 @command{nvprof}, that @code{acc_get_device_type}
5760 (@ref{acc_get_device_type})) may be called from
5761 @code{acc_ev_device_init_start}, @code{acc_ev_device_init_end}
5764 We're not yet implementing initialization via a
5765 @code{acc_register_library} function that is either statically linked
5766 in, or dynamically via @env{LD_PRELOAD}.
5767 Initialization via @code{acc_register_library} functions dynamically
5768 loaded via the @env{ACC_PROFLIB} environment variable does work, as
5769 does directly calling @code{acc_prof_register},
5770 @code{acc_prof_unregister}, @code{acc_prof_lookup}.
5772 As currently there are no inquiry functions defined, calls to
5773 @code{acc_prof_lookup} always returns @code{NULL}.
5775 There aren't separate @emph{start}, @emph{stop} events defined for the
5776 event types @code{acc_ev_create}, @code{acc_ev_delete},
5777 @code{acc_ev_alloc}, @code{acc_ev_free}. It's not clear if these
5778 should be triggered before or after the actual device-specific call is
5779 made. We trigger them after.
5781 Remarks about data provided to callbacks:
5785 @item @code{acc_prof_info.event_type}
5786 It's not clear if for @emph{nested} event callbacks (for example,
5787 @code{acc_ev_enqueue_launch_start} as part of a parent compute
5788 construct), this should be set for the nested event
5789 (@code{acc_ev_enqueue_launch_start}), or if the value of the parent
5790 construct should remain (@code{acc_ev_compute_construct_start}). In
5791 this implementation, the value generally corresponds to the
5792 innermost nested event type.
5794 @item @code{acc_prof_info.device_type}
5798 For @code{acc_ev_compute_construct_start}, and in presence of an
5799 @code{if} clause with @emph{false} argument, this still refers to
5800 the offloading device type.
5801 It's not clear if that's the expected behavior.
5804 Complementary to the item before, for
5805 @code{acc_ev_compute_construct_end}, this is set to
5806 @code{acc_device_host} in presence of an @code{if} clause with
5807 @emph{false} argument.
5808 It's not clear if that's the expected behavior.
5812 @item @code{acc_prof_info.thread_id}
5813 Always @code{-1}; not yet implemented.
5815 @item @code{acc_prof_info.async}
5819 Not yet implemented correctly for
5820 @code{acc_ev_compute_construct_start}.
5823 In a compute construct, for host-fallback
5824 execution/@code{acc_device_host} it always is
5825 @code{acc_async_sync}.
5826 It is unclear if that is the expected behavior.
5829 For @code{acc_ev_device_init_start} and @code{acc_ev_device_init_end},
5830 it will always be @code{acc_async_sync}.
5831 It is unclear if that is the expected behavior.
5835 @item @code{acc_prof_info.async_queue}
5836 There is no @cite{limited number of asynchronous queues} in libgomp.
5837 This always has the same value as @code{acc_prof_info.async}.
5839 @item @code{acc_prof_info.src_file}
5840 Always @code{NULL}; not yet implemented.
5842 @item @code{acc_prof_info.func_name}
5843 Always @code{NULL}; not yet implemented.
5845 @item @code{acc_prof_info.line_no}
5846 Always @code{-1}; not yet implemented.
5848 @item @code{acc_prof_info.end_line_no}
5849 Always @code{-1}; not yet implemented.
5851 @item @code{acc_prof_info.func_line_no}
5852 Always @code{-1}; not yet implemented.
5854 @item @code{acc_prof_info.func_end_line_no}
5855 Always @code{-1}; not yet implemented.
5857 @item @code{acc_event_info.event_type}, @code{acc_event_info.*.event_type}
5858 Relating to @code{acc_prof_info.event_type} discussed above, in this
5859 implementation, this will always be the same value as
5860 @code{acc_prof_info.event_type}.
5862 @item @code{acc_event_info.*.parent_construct}
5866 Will be @code{acc_construct_parallel} for all OpenACC compute
5867 constructs as well as many OpenACC Runtime API calls; should be the
5868 one matching the actual construct, or
5869 @code{acc_construct_runtime_api}, respectively.
5872 Will be @code{acc_construct_enter_data} or
5873 @code{acc_construct_exit_data} when processing variable mappings
5874 specified in OpenACC @emph{declare} directives; should be
5875 @code{acc_construct_declare}.
5878 For implicit @code{acc_ev_device_init_start},
5879 @code{acc_ev_device_init_end}, and explicit as well as implicit
5880 @code{acc_ev_alloc}, @code{acc_ev_free},
5881 @code{acc_ev_enqueue_upload_start}, @code{acc_ev_enqueue_upload_end},
5882 @code{acc_ev_enqueue_download_start}, and
5883 @code{acc_ev_enqueue_download_end}, will be
5884 @code{acc_construct_parallel}; should reflect the real parent
5889 @item @code{acc_event_info.*.implicit}
5890 For @code{acc_ev_alloc}, @code{acc_ev_free},
5891 @code{acc_ev_enqueue_upload_start}, @code{acc_ev_enqueue_upload_end},
5892 @code{acc_ev_enqueue_download_start}, and
5893 @code{acc_ev_enqueue_download_end}, this currently will be @code{1}
5894 also for explicit usage.
5896 @item @code{acc_event_info.data_event.var_name}
5897 Always @code{NULL}; not yet implemented.
5899 @item @code{acc_event_info.data_event.host_ptr}
5900 For @code{acc_ev_alloc}, and @code{acc_ev_free}, this is always
5903 @item @code{typedef union acc_api_info}
5904 @dots{} as printed in @cite{5.2.3. Third Argument: API-Specific
5905 Information}. This should obviously be @code{typedef @emph{struct}
5908 @item @code{acc_api_info.device_api}
5909 Possibly not yet implemented correctly for
5910 @code{acc_ev_compute_construct_start},
5911 @code{acc_ev_device_init_start}, @code{acc_ev_device_init_end}:
5912 will always be @code{acc_device_api_none} for these event types.
5913 For @code{acc_ev_enter_data_start}, it will be
5914 @code{acc_device_api_none} in some cases.
5916 @item @code{acc_api_info.device_type}
5917 Always the same as @code{acc_prof_info.device_type}.
5919 @item @code{acc_api_info.vendor}
5920 Always @code{-1}; not yet implemented.
5922 @item @code{acc_api_info.device_handle}
5923 Always @code{NULL}; not yet implemented.
5925 @item @code{acc_api_info.context_handle}
5926 Always @code{NULL}; not yet implemented.
5928 @item @code{acc_api_info.async_handle}
5929 Always @code{NULL}; not yet implemented.
5933 Remarks about certain event types:
5937 @item @code{acc_ev_device_init_start}, @code{acc_ev_device_init_end}
5941 @c See 'DEVICE_INIT_INSIDE_COMPUTE_CONSTRUCT' in
5942 @c 'libgomp.oacc-c-c++-common/acc_prof-kernels-1.c',
5943 @c 'libgomp.oacc-c-c++-common/acc_prof-parallel-1.c'.
5944 When a compute construct triggers implicit
5945 @code{acc_ev_device_init_start} and @code{acc_ev_device_init_end}
5946 events, they currently aren't @emph{nested within} the corresponding
5947 @code{acc_ev_compute_construct_start} and
5948 @code{acc_ev_compute_construct_end}, but they're currently observed
5949 @emph{before} @code{acc_ev_compute_construct_start}.
5950 It's not clear what to do: the standard asks us provide a lot of
5951 details to the @code{acc_ev_compute_construct_start} callback, without
5952 (implicitly) initializing a device before?
5955 Callbacks for these event types will not be invoked for calls to the
5956 @code{acc_set_device_type} and @code{acc_set_device_num} functions.
5957 It's not clear if they should be.
5961 @item @code{acc_ev_enter_data_start}, @code{acc_ev_enter_data_end}, @code{acc_ev_exit_data_start}, @code{acc_ev_exit_data_end}
5965 Callbacks for these event types will also be invoked for OpenACC
5966 @emph{host_data} constructs.
5967 It's not clear if they should be.
5970 Callbacks for these event types will also be invoked when processing
5971 variable mappings specified in OpenACC @emph{declare} directives.
5972 It's not clear if they should be.
5978 Callbacks for the following event types will be invoked, but dispatch
5979 and information provided therein has not yet been thoroughly reviewed:
5982 @item @code{acc_ev_alloc}
5983 @item @code{acc_ev_free}
5984 @item @code{acc_ev_update_start}, @code{acc_ev_update_end}
5985 @item @code{acc_ev_enqueue_upload_start}, @code{acc_ev_enqueue_upload_end}
5986 @item @code{acc_ev_enqueue_download_start}, @code{acc_ev_enqueue_download_end}
5989 During device initialization, and finalization, respectively,
5990 callbacks for the following event types will not yet be invoked:
5993 @item @code{acc_ev_alloc}
5994 @item @code{acc_ev_free}
5997 Callbacks for the following event types have not yet been implemented,
5998 so currently won't be invoked:
6001 @item @code{acc_ev_device_shutdown_start}, @code{acc_ev_device_shutdown_end}
6002 @item @code{acc_ev_runtime_shutdown}
6003 @item @code{acc_ev_create}, @code{acc_ev_delete}
6004 @item @code{acc_ev_wait_start}, @code{acc_ev_wait_end}
6007 For the following runtime library functions, not all expected
6008 callbacks will be invoked (mostly concerning implicit device
6012 @item @code{acc_get_num_devices}
6013 @item @code{acc_set_device_type}
6014 @item @code{acc_get_device_type}
6015 @item @code{acc_set_device_num}
6016 @item @code{acc_get_device_num}
6017 @item @code{acc_init}
6018 @item @code{acc_shutdown}
6021 Aside from implicit device initialization, for the following runtime
6022 library functions, no callbacks will be invoked for shared-memory
6023 offloading devices (it's not clear if they should be):
6026 @item @code{acc_malloc}
6027 @item @code{acc_free}
6028 @item @code{acc_copyin}, @code{acc_present_or_copyin}, @code{acc_copyin_async}
6029 @item @code{acc_create}, @code{acc_present_or_create}, @code{acc_create_async}
6030 @item @code{acc_copyout}, @code{acc_copyout_async}, @code{acc_copyout_finalize}, @code{acc_copyout_finalize_async}
6031 @item @code{acc_delete}, @code{acc_delete_async}, @code{acc_delete_finalize}, @code{acc_delete_finalize_async}
6032 @item @code{acc_update_device}, @code{acc_update_device_async}
6033 @item @code{acc_update_self}, @code{acc_update_self_async}
6034 @item @code{acc_map_data}, @code{acc_unmap_data}
6035 @item @code{acc_memcpy_to_device}, @code{acc_memcpy_to_device_async}
6036 @item @code{acc_memcpy_from_device}, @code{acc_memcpy_from_device_async}
6039 @c ---------------------------------------------------------------------
6040 @c OpenMP-Implementation Specifics
6041 @c ---------------------------------------------------------------------
6043 @node OpenMP-Implementation Specifics
6044 @chapter OpenMP-Implementation Specifics
6047 * Implementation-defined ICV Initialization::
6048 * OpenMP Context Selectors::
6049 * Memory allocation::
6052 @node Implementation-defined ICV Initialization
6053 @section Implementation-defined ICV Initialization
6054 @cindex Implementation specific setting
6056 @multitable @columnfractions .30 .70
6057 @item @var{affinity-format-var} @tab See @ref{OMP_AFFINITY_FORMAT}.
6058 @item @var{def-allocator-var} @tab See @ref{OMP_ALLOCATOR}.
6059 @item @var{max-active-levels-var} @tab See @ref{OMP_MAX_ACTIVE_LEVELS}.
6060 @item @var{dyn-var} @tab See @ref{OMP_DYNAMIC}.
6061 @item @var{nthreads-var} @tab See @ref{OMP_NUM_THREADS}.
6062 @item @var{num-devices-var} @tab Number of non-host devices found
6063 by GCC's run-time library
6064 @item @var{num-procs-var} @tab The number of CPU cores on the
6065 initial device, except that affinity settings might lead to a
6066 smaller number. On non-host devices, the value of the
6067 @var{nthreads-var} ICV.
6068 @item @var{place-partition-var} @tab See @ref{OMP_PLACES}.
6069 @item @var{run-sched-var} @tab See @ref{OMP_SCHEDULE}.
6070 @item @var{stacksize-var} @tab See @ref{OMP_STACKSIZE}.
6071 @item @var{thread-limit-var} @tab See @ref{OMP_TEAMS_THREAD_LIMIT}
6072 @item @var{wait-policy-var} @tab See @ref{OMP_WAIT_POLICY} and
6073 @ref{GOMP_SPINCOUNT}
6076 @node OpenMP Context Selectors
6077 @section OpenMP Context Selectors
6079 @code{vendor} is always @code{gnu}. References are to the GCC manual.
6081 @c NOTE: Only the following selectors have been implemented. To add
6082 @c additional traits for target architecture, TARGET_OMP_DEVICE_KIND_ARCH_ISA
6083 @c has to be implemented; cf. also PR target/105640.
6084 @c For offload devices, add *additionally* gcc/config/*/t-omp-device.
6086 For the host compiler, @code{kind} always matches @code{host}; for the
6087 offloading architectures AMD GCN and Nvidia PTX, @code{kind} always matches
6088 @code{gpu}. For the x86 family of computers, AMD GCN and Nvidia PTX
6089 the following traits are supported in addition; while OpenMP is supported
6090 on more architectures, GCC currently does not match any @code{arch} or
6091 @code{isa} traits for those.
6093 @multitable @columnfractions .65 .30
6094 @headitem @code{arch} @tab @code{isa}
6095 @item @code{x86}, @code{x86_64}, @code{i386}, @code{i486},
6096 @code{i586}, @code{i686}, @code{ia32}
6097 @tab See @code{-m...} flags in ``x86 Options'' (without @code{-m})
6098 @item @code{amdgcn}, @code{gcn}
6099 @tab See @code{-march=} in ``AMD GCN Options''@footnote{Additionally,
6100 @code{gfx803} is supported as an alias for @code{fiji}.}
6102 @tab See @code{-march=} in ``Nvidia PTX Options''
6105 @node Memory allocation
6106 @section Memory allocation
6108 The description below applies to:
6111 @item Explicit use of the OpenMP API routines, see
6112 @ref{Memory Management Routines}.
6113 @item The @code{allocate} clause, except when the @code{allocator} modifier is a
6114 constant expression with value @code{omp_default_mem_alloc} and no
6115 @code{align} modifier has been specified. (In that case, the normal
6116 @code{malloc} allocation is used.)
6117 @item Using the @code{allocate} directive for automatic/stack variables, except
6118 when the @code{allocator} clause is a constant expression with value
6119 @code{omp_default_mem_alloc} and no @code{align} clause has been
6120 specified. (In that case, the normal allocation is used: stack allocation
6121 and, sometimes for Fortran, also @code{malloc} [depending on flags such as
6122 @option{-fstack-arrays}].)
6123 @item Using the @code{allocate} directive for variable in static memory is
6124 currently not supported (compile time error).
6125 @item In Fortran, the @code{allocators} directive and the executable
6126 @code{allocate} directive for Fortran pointers and allocatables is
6127 supported, but requires that files containing those directives has to be
6128 compiled with @option{-fopenmp-allocators}. Additionally, all files that
6129 might explicitly or implicitly deallocate memory allocated that way must
6130 also be compiled with that option.
6133 For the available predefined allocators and, as applicable, their associated
6134 predefined memory spaces and for the available traits and their default values,
6135 see @ref{OMP_ALLOCATOR}. Predefined allocators without an associated memory
6136 space use the @code{omp_default_mem_space} memory space.
6138 For the memory spaces, the following applies:
6140 @item @code{omp_default_mem_space} is supported
6141 @item @code{omp_const_mem_space} maps to @code{omp_default_mem_space}
6142 @item @code{omp_low_lat_mem_space} is only available on supported devices,
6143 and maps to @code{omp_default_mem_space} otherwise.
6144 @item @code{omp_large_cap_mem_space} maps to @code{omp_default_mem_space},
6145 unless the memkind library is available
6146 @item @code{omp_high_bw_mem_space} maps to @code{omp_default_mem_space},
6147 unless the memkind library is available
6150 On Linux systems, where the @uref{https://github.com/memkind/memkind, memkind
6151 library} (@code{libmemkind.so.0}) is available at runtime, it is used when
6152 creating memory allocators requesting
6155 @item the memory space @code{omp_high_bw_mem_space}
6156 @item the memory space @code{omp_large_cap_mem_space}
6157 @item the @code{partition} trait @code{interleaved}; note that for
6158 @code{omp_large_cap_mem_space} the allocation will not be interleaved
6161 On Linux systems, where the @uref{https://github.com/numactl/numactl, numa
6162 library} (@code{libnuma.so.1}) is available at runtime, it used when creating
6163 memory allocators requesting
6166 @item the @code{partition} trait @code{nearest}, except when both the
6167 libmemkind library is available and the memory space is either
6168 @code{omp_large_cap_mem_space} or @code{omp_high_bw_mem_space}
6171 Note that the numa library will round up the allocation size to a multiple of
6172 the system page size; therefore, consider using it only with large data or
6173 by sharing allocations via the @code{pool_size} trait. Furthermore, the Linux
6174 kernel does not guarantee that an allocation will always be on the nearest NUMA
6175 node nor that after reallocation the same node will be used. Note additionally
6176 that, on Linux, the default setting of the memory placement policy is to use the
6177 current node; therefore, unless the memory placement policy has been overridden,
6178 the @code{partition} trait @code{environment} (the default) will be effectively
6179 a @code{nearest} allocation.
6181 Additional notes regarding the traits:
6183 @item The @code{pinned} trait is supported on Linux hosts, but is subject to
6184 the OS @code{ulimit}/@code{rlimit} locked memory settings.
6185 @item The default for the @code{pool_size} trait is no pool and for every
6186 (re)allocation the associated library routine is called, which might
6187 internally use a memory pool.
6188 @item For the @code{partition} trait, the partition part size will be the same
6189 as the requested size (i.e. @code{interleaved} or @code{blocked} has no
6190 effect), except for @code{interleaved} when the memkind library is
6191 available. Furthermore, for @code{nearest} and unless the numa library
6192 is available, the memory might not be on the same NUMA node as thread
6193 that allocated the memory; on Linux, this is in particular the case when
6194 the memory placement policy is set to preferred.
6195 @item The @code{access} trait has no effect such that memory is always
6196 accessible by all threads.
6197 @item The @code{sync_hint} trait has no effect.
6201 @ref{Offload-Target Specifics}
6203 @c ---------------------------------------------------------------------
6204 @c Offload-Target Specifics
6205 @c ---------------------------------------------------------------------
6207 @node Offload-Target Specifics
6208 @chapter Offload-Target Specifics
6210 The following sections present notes on the offload-target specifics
6218 @section AMD Radeon (GCN)
6220 On the hardware side, there is the hierarchy (fine to coarse):
6222 @item work item (thread)
6225 @item compute unit (CU)
6228 All OpenMP and OpenACC levels are used, i.e.
6230 @item OpenMP's simd and OpenACC's vector map to work items (thread)
6231 @item OpenMP's threads (``parallel'') and OpenACC's workers map
6233 @item OpenMP's teams and OpenACC's gang use a threadpool with the
6234 size of the number of teams or gangs, respectively.
6239 @item Number of teams is the specified @code{num_teams} (OpenMP) or
6240 @code{num_gangs} (OpenACC) or otherwise the number of CU. It is limited
6241 by two times the number of CU.
6242 @item Number of wavefronts is 4 for gfx900 and 16 otherwise;
6243 @code{num_threads} (OpenMP) and @code{num_workers} (OpenACC)
6244 overrides this if smaller.
6245 @item The wavefront has 102 scalars and 64 vectors
6246 @item Number of workitems is always 64
6247 @item The hardware permits maximally 40 workgroups/CU and
6248 16 wavefronts/workgroup up to a limit of 40 wavefronts in total per CU.
6249 @item 80 scalars registers and 24 vector registers in non-kernel functions
6250 (the chosen procedure-calling API).
6251 @item For the kernel itself: as many as register pressure demands (number of
6252 teams and number of threads, scaled down if registers are exhausted)
6255 The implementation remark:
6257 @item I/O within OpenMP target regions and OpenACC parallel/kernels is supported
6258 using the C library @code{printf} functions and the Fortran
6259 @code{print}/@code{write} statements.
6260 @item Reverse offload regions (i.e. @code{target} regions with
6261 @code{device(ancestor:1)}) are processed serially per @code{target} region
6262 such that the next reverse offload region is only executed after the previous
6264 @item OpenMP code that has a @code{requires} directive with
6265 @code{unified_shared_memory} will remove any GCN device from the list of
6266 available devices (``host fallback'').
6267 @item The available stack size can be changed using the @code{GCN_STACK_SIZE}
6268 environment variable; the default is 32 kiB per thread.
6269 @item Low-latency memory (@code{omp_low_lat_mem_space}) is supported when the
6270 the @code{access} trait is set to @code{cgroup}. The default pool size
6271 is automatically scaled to share the 64 kiB LDS memory between the number
6272 of teams configured to run on each compute-unit, but may be adjusted at
6273 runtime by setting environment variable
6274 @code{GOMP_GCN_LOWLAT_POOL=@var{bytes}}.
6275 @item @code{omp_low_lat_mem_alloc} cannot be used with true low-latency memory
6276 because the definition implies the @code{omp_atv_all} trait; main
6277 graphics memory is used instead.
6278 @item @code{omp_cgroup_mem_alloc}, @code{omp_pteam_mem_alloc}, and
6279 @code{omp_thread_mem_alloc}, all use low-latency memory as first
6280 preference, and fall back to main graphics memory when the low-latency
6289 On the hardware side, there is the hierarchy (fine to coarse):
6294 @item streaming multiprocessor
6297 All OpenMP and OpenACC levels are used, i.e.
6299 @item OpenMP's simd and OpenACC's vector map to threads
6300 @item OpenMP's threads (``parallel'') and OpenACC's workers map to warps
6301 @item OpenMP's teams and OpenACC's gang use a threadpool with the
6302 size of the number of teams or gangs, respectively.
6307 @item The @code{warp_size} is always 32
6308 @item CUDA kernel launched: @code{dim=@{#teams,1,1@}, blocks=@{#threads,warp_size,1@}}.
6309 @item The number of teams is limited by the number of blocks the device can
6310 host simultaneously.
6313 Additional information can be obtained by setting the environment variable to
6314 @code{GOMP_DEBUG=1} (very verbose; grep for @code{kernel.*launch} for launch
6317 GCC generates generic PTX ISA code, which is just-in-time compiled by CUDA,
6318 which caches the JIT in the user's directory (see CUDA documentation; can be
6319 tuned by the environment variables @code{CUDA_CACHE_@{DISABLE,MAXSIZE,PATH@}}.
6321 Note: While PTX ISA is generic, the @code{-mptx=} and @code{-march=} commandline
6322 options still affect the used PTX ISA code and, thus, the requirements on
6323 CUDA version and hardware.
6325 The implementation remark:
6327 @item I/O within OpenMP target regions and OpenACC parallel/kernels is supported
6328 using the C library @code{printf} functions. Note that the Fortran
6329 @code{print}/@code{write} statements are not supported, yet.
6330 @item Compilation OpenMP code that contains @code{requires reverse_offload}
6331 requires at least @code{-march=sm_35}, compiling for @code{-march=sm_30}
6333 @item For code containing reverse offload (i.e. @code{target} regions with
6334 @code{device(ancestor:1)}), there is a slight performance penalty
6335 for @emph{all} target regions, consisting mostly of shutdown delay
6336 Per device, reverse offload regions are processed serially such that
6337 the next reverse offload region is only executed after the previous
6339 @item OpenMP code that has a @code{requires} directive with
6340 @code{unified_shared_memory} will remove any nvptx device from the
6341 list of available devices (``host fallback'').
6342 @item The default per-warp stack size is 128 kiB; see also @code{-msoft-stack}
6344 @item The OpenMP routines @code{omp_target_memcpy_rect} and
6345 @code{omp_target_memcpy_rect_async} and the @code{target update}
6346 directive for non-contiguous list items will use the 2D and 3D
6347 memory-copy functions of the CUDA library. Higher dimensions will
6348 call those functions in a loop and are therefore supported.
6349 @item Low-latency memory (@code{omp_low_lat_mem_space}) is supported when the
6350 the @code{access} trait is set to @code{cgroup}, the ISA is at least
6351 @code{sm_53}, and the PTX version is at least 4.1. The default pool size
6352 is 8 kiB per team, but may be adjusted at runtime by setting environment
6353 variable @code{GOMP_NVPTX_LOWLAT_POOL=@var{bytes}}. The maximum value is
6354 limited by the available hardware, and care should be taken that the
6355 selected pool size does not unduly limit the number of teams that can
6357 @item @code{omp_low_lat_mem_alloc} cannot be used with true low-latency memory
6358 because the definition implies the @code{omp_atv_all} trait; main
6359 graphics memory is used instead.
6360 @item @code{omp_cgroup_mem_alloc}, @code{omp_pteam_mem_alloc}, and
6361 @code{omp_thread_mem_alloc}, all use low-latency memory as first
6362 preference, and fall back to main graphics memory when the low-latency
6367 @c ---------------------------------------------------------------------
6369 @c ---------------------------------------------------------------------
6371 @node The libgomp ABI
6372 @chapter The libgomp ABI
6374 The following sections present notes on the external ABI as
6375 presented by libgomp. Only maintainers should need them.
6378 * Implementing MASTER construct::
6379 * Implementing CRITICAL construct::
6380 * Implementing ATOMIC construct::
6381 * Implementing FLUSH construct::
6382 * Implementing BARRIER construct::
6383 * Implementing THREADPRIVATE construct::
6384 * Implementing PRIVATE clause::
6385 * Implementing FIRSTPRIVATE LASTPRIVATE COPYIN and COPYPRIVATE clauses::
6386 * Implementing REDUCTION clause::
6387 * Implementing PARALLEL construct::
6388 * Implementing FOR construct::
6389 * Implementing ORDERED construct::
6390 * Implementing SECTIONS construct::
6391 * Implementing SINGLE construct::
6392 * Implementing OpenACC's PARALLEL construct::
6396 @node Implementing MASTER construct
6397 @section Implementing MASTER construct
6400 if (omp_get_thread_num () == 0)
6404 Alternately, we generate two copies of the parallel subfunction
6405 and only include this in the version run by the primary thread.
6406 Surely this is not worthwhile though...
6410 @node Implementing CRITICAL construct
6411 @section Implementing CRITICAL construct
6413 Without a specified name,
6416 void GOMP_critical_start (void);
6417 void GOMP_critical_end (void);
6420 so that we don't get COPY relocations from libgomp to the main
6423 With a specified name, use omp_set_lock and omp_unset_lock with
6424 name being transformed into a variable declared like
6427 omp_lock_t gomp_critical_user_<name> __attribute__((common))
6430 Ideally the ABI would specify that all zero is a valid unlocked
6431 state, and so we wouldn't need to initialize this at
6436 @node Implementing ATOMIC construct
6437 @section Implementing ATOMIC construct
6439 The target should implement the @code{__sync} builtins.
6441 Failing that we could add
6444 void GOMP_atomic_enter (void)
6445 void GOMP_atomic_exit (void)
6448 which reuses the regular lock code, but with yet another lock
6449 object private to the library.
6453 @node Implementing FLUSH construct
6454 @section Implementing FLUSH construct
6456 Expands to the @code{__sync_synchronize} builtin.
6460 @node Implementing BARRIER construct
6461 @section Implementing BARRIER construct
6464 void GOMP_barrier (void)
6468 @node Implementing THREADPRIVATE construct
6469 @section Implementing THREADPRIVATE construct
6471 In _most_ cases we can map this directly to @code{__thread}. Except
6472 that OMP allows constructors for C++ objects. We can either
6473 refuse to support this (how often is it used?) or we can
6474 implement something akin to .ctors.
6476 Even more ideally, this ctor feature is handled by extensions
6477 to the main pthreads library. Failing that, we can have a set
6478 of entry points to register ctor functions to be called.
6482 @node Implementing PRIVATE clause
6483 @section Implementing PRIVATE clause
6485 In association with a PARALLEL, or within the lexical extent
6486 of a PARALLEL block, the variable becomes a local variable in
6487 the parallel subfunction.
6489 In association with FOR or SECTIONS blocks, create a new
6490 automatic variable within the current function. This preserves
6491 the semantic of new variable creation.
6495 @node Implementing FIRSTPRIVATE LASTPRIVATE COPYIN and COPYPRIVATE clauses
6496 @section Implementing FIRSTPRIVATE LASTPRIVATE COPYIN and COPYPRIVATE clauses
6498 This seems simple enough for PARALLEL blocks. Create a private
6499 struct for communicating between the parent and subfunction.
6500 In the parent, copy in values for scalar and "small" structs;
6501 copy in addresses for others TREE_ADDRESSABLE types. In the
6502 subfunction, copy the value into the local variable.
6504 It is not clear what to do with bare FOR or SECTION blocks.
6505 The only thing I can figure is that we do something like:
6508 #pragma omp for firstprivate(x) lastprivate(y)
6509 for (int i = 0; i < n; ++i)
6526 where the "x=x" and "y=y" assignments actually have different
6527 uids for the two variables, i.e. not something you could write
6528 directly in C. Presumably this only makes sense if the "outer"
6529 x and y are global variables.
6531 COPYPRIVATE would work the same way, except the structure
6532 broadcast would have to happen via SINGLE machinery instead.
6536 @node Implementing REDUCTION clause
6537 @section Implementing REDUCTION clause
6539 The private struct mentioned in the previous section should have
6540 a pointer to an array of the type of the variable, indexed by the
6541 thread's @var{team_id}. The thread stores its final value into the
6542 array, and after the barrier, the primary thread iterates over the
6543 array to collect the values.
6546 @node Implementing PARALLEL construct
6547 @section Implementing PARALLEL construct
6550 #pragma omp parallel
6559 void subfunction (void *data)
6566 GOMP_parallel_start (subfunction, &data, num_threads);
6567 subfunction (&data);
6568 GOMP_parallel_end ();
6572 void GOMP_parallel_start (void (*fn)(void *), void *data, unsigned num_threads)
6575 The @var{FN} argument is the subfunction to be run in parallel.
6577 The @var{DATA} argument is a pointer to a structure used to
6578 communicate data in and out of the subfunction, as discussed
6579 above with respect to FIRSTPRIVATE et al.
6581 The @var{NUM_THREADS} argument is 1 if an IF clause is present
6582 and false, or the value of the NUM_THREADS clause, if
6585 The function needs to create the appropriate number of
6586 threads and/or launch them from the dock. It needs to
6587 create the team structure and assign team ids.
6590 void GOMP_parallel_end (void)
6593 Tears down the team and returns us to the previous @code{omp_in_parallel()} state.
6597 @node Implementing FOR construct
6598 @section Implementing FOR construct
6601 #pragma omp parallel for
6602 for (i = lb; i <= ub; i++)
6609 void subfunction (void *data)
6612 while (GOMP_loop_static_next (&_s0, &_e0))
6615 for (i = _s0; i < _e1; i++)
6618 GOMP_loop_end_nowait ();
6621 GOMP_parallel_loop_static (subfunction, NULL, 0, lb, ub+1, 1, 0);
6623 GOMP_parallel_end ();
6627 #pragma omp for schedule(runtime)
6628 for (i = 0; i < n; i++)
6637 if (GOMP_loop_runtime_start (0, n, 1, &_s0, &_e0))
6640 for (i = _s0, i < _e0; i++)
6642 @} while (GOMP_loop_runtime_next (&_s0, _&e0));
6647 Note that while it looks like there is trickiness to propagating
6648 a non-constant STEP, there isn't really. We're explicitly allowed
6649 to evaluate it as many times as we want, and any variables involved
6650 should automatically be handled as PRIVATE or SHARED like any other
6651 variables. So the expression should remain evaluable in the
6652 subfunction. We can also pull it into a local variable if we like,
6653 but since its supposed to remain unchanged, we can also not if we like.
6655 If we have SCHEDULE(STATIC), and no ORDERED, then we ought to be
6656 able to get away with no work-sharing context at all, since we can
6657 simply perform the arithmetic directly in each thread to divide up
6658 the iterations. Which would mean that we wouldn't need to call any
6661 There are separate routines for handling loops with an ORDERED
6662 clause. Bookkeeping for that is non-trivial...
6666 @node Implementing ORDERED construct
6667 @section Implementing ORDERED construct
6670 void GOMP_ordered_start (void)
6671 void GOMP_ordered_end (void)
6676 @node Implementing SECTIONS construct
6677 @section Implementing SECTIONS construct
6682 #pragma omp sections
6696 for (i = GOMP_sections_start (3); i != 0; i = GOMP_sections_next ())
6713 @node Implementing SINGLE construct
6714 @section Implementing SINGLE construct
6728 if (GOMP_single_start ())
6736 #pragma omp single copyprivate(x)
6743 datap = GOMP_single_copy_start ();
6748 GOMP_single_copy_end (&data);
6757 @node Implementing OpenACC's PARALLEL construct
6758 @section Implementing OpenACC's PARALLEL construct
6761 void GOACC_parallel ()
6766 @c ---------------------------------------------------------------------
6768 @c ---------------------------------------------------------------------
6770 @node Reporting Bugs
6771 @chapter Reporting Bugs
6773 Bugs in the GNU Offloading and Multi Processing Runtime Library should
6774 be reported via @uref{https://gcc.gnu.org/bugzilla/, Bugzilla}. Please add
6775 "openacc", or "openmp", or both to the keywords field in the bug
6776 report, as appropriate.
6780 @c ---------------------------------------------------------------------
6781 @c GNU General Public License
6782 @c ---------------------------------------------------------------------
6784 @include gpl_v3.texi
6788 @c ---------------------------------------------------------------------
6789 @c GNU Free Documentation License
6790 @c ---------------------------------------------------------------------
6796 @c ---------------------------------------------------------------------
6797 @c Funding Free Software
6798 @c ---------------------------------------------------------------------
6800 @include funding.texi
6802 @c ---------------------------------------------------------------------
6804 @c ---------------------------------------------------------------------
6807 @unnumbered Library Index