1 \input texinfo @c -*-texinfo-*-
4 @setfilename libgomp.info
10 Copyright @copyright{} 2006-2020 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@*
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 Runtime Library Routines: Runtime Library Routines.
99 The OpenMP runtime application programming
101 * OpenMP Environment Variables: Environment Variables.
102 Influencing OpenMP runtime behavior with
103 environment variables.
104 * Enabling OpenACC:: How to enable OpenACC for your
106 * OpenACC Runtime Library Routines:: The OpenACC runtime application
107 programming interface.
108 * OpenACC Environment Variables:: Influencing OpenACC runtime behavior with
109 environment variables.
110 * CUDA Streams Usage:: Notes on the implementation of
111 asynchronous operations.
112 * OpenACC Library Interoperability:: OpenACC library interoperability with the
113 NVIDIA CUBLAS library.
114 * OpenACC Profiling Interface::
115 * The libgomp ABI:: Notes on the external ABI presented by libgomp.
116 * Reporting Bugs:: How to report bugs in the GNU Offloading and
117 Multi Processing Runtime Library.
118 * Copying:: GNU general public license says
119 how you can copy and share libgomp.
120 * GNU Free Documentation License::
121 How you can copy and share this manual.
122 * Funding:: How to help assure continued work for free
124 * Library Index:: Index of this documentation.
128 @c ---------------------------------------------------------------------
130 @c ---------------------------------------------------------------------
132 @node Enabling OpenMP
133 @chapter Enabling OpenMP
135 To activate the OpenMP extensions for C/C++ and Fortran, the compile-time
136 flag @command{-fopenmp} must be specified. This enables the OpenMP directive
137 @code{#pragma omp} in C/C++ and @code{!$omp} directives in free form,
138 @code{c$omp}, @code{*$omp} and @code{!$omp} directives in fixed form,
139 @code{!$} conditional compilation sentinels in free form and @code{c$},
140 @code{*$} and @code{!$} sentinels in fixed form, for Fortran. The flag also
141 arranges for automatic linking of the OpenMP runtime library
142 (@ref{Runtime Library Routines}).
144 A complete description of all OpenMP directives accepted may be found in
145 the @uref{https://www.openmp.org, OpenMP Application Program Interface} manual,
149 @c ---------------------------------------------------------------------
150 @c OpenMP Runtime Library Routines
151 @c ---------------------------------------------------------------------
153 @node Runtime Library Routines
154 @chapter OpenMP Runtime Library Routines
156 The runtime routines described here are defined by Section 3 of the OpenMP
157 specification in version 4.5. The routines are structured in following
161 Control threads, processors and the parallel environment. They have C
162 linkage, and do not throw exceptions.
164 * omp_get_active_level:: Number of active parallel regions
165 * omp_get_ancestor_thread_num:: Ancestor thread ID
166 * omp_get_cancellation:: Whether cancellation support is enabled
167 * omp_get_default_device:: Get the default device for target regions
168 * omp_get_dynamic:: Dynamic teams setting
169 * omp_get_level:: Number of parallel regions
170 * omp_get_max_active_levels:: Maximum number of active regions
171 * omp_get_max_task_priority:: Maximum task priority value that can be set
172 * omp_get_max_threads:: Maximum number of threads of parallel region
173 * omp_get_nested:: Nested parallel regions
174 * omp_get_num_devices:: Number of target devices
175 * omp_get_num_procs:: Number of processors online
176 * omp_get_num_teams:: Number of teams
177 * omp_get_num_threads:: Size of the active team
178 * omp_get_proc_bind:: Whether theads may be moved between CPUs
179 * omp_get_schedule:: Obtain the runtime scheduling method
180 * omp_get_team_num:: Get team number
181 * omp_get_team_size:: Number of threads in a team
182 * omp_get_thread_limit:: Maximum number of threads
183 * omp_get_thread_num:: Current thread ID
184 * omp_in_parallel:: Whether a parallel region is active
185 * omp_in_final:: Whether in final or included task region
186 * omp_is_initial_device:: Whether executing on the host device
187 * omp_set_default_device:: Set the default device for target regions
188 * omp_set_dynamic:: Enable/disable dynamic teams
189 * omp_set_max_active_levels:: Limits the number of active parallel regions
190 * omp_set_nested:: Enable/disable nested parallel regions
191 * omp_set_num_threads:: Set upper team size limit
192 * omp_set_schedule:: Set the runtime scheduling method
194 Initialize, set, test, unset and destroy simple and nested locks.
196 * omp_init_lock:: Initialize simple lock
197 * omp_set_lock:: Wait for and set simple lock
198 * omp_test_lock:: Test and set simple lock if available
199 * omp_unset_lock:: Unset simple lock
200 * omp_destroy_lock:: Destroy simple lock
201 * omp_init_nest_lock:: Initialize nested lock
202 * omp_set_nest_lock:: Wait for and set simple lock
203 * omp_test_nest_lock:: Test and set nested lock if available
204 * omp_unset_nest_lock:: Unset nested lock
205 * omp_destroy_nest_lock:: Destroy nested lock
207 Portable, thread-based, wall clock timer.
209 * omp_get_wtick:: Get timer precision.
210 * omp_get_wtime:: Elapsed wall clock time.
215 @node omp_get_active_level
216 @section @code{omp_get_active_level} -- Number of parallel regions
218 @item @emph{Description}:
219 This function returns the nesting level for the active parallel blocks,
220 which enclose the calling call.
223 @multitable @columnfractions .20 .80
224 @item @emph{Prototype}: @tab @code{int omp_get_active_level(void);}
227 @item @emph{Fortran}:
228 @multitable @columnfractions .20 .80
229 @item @emph{Interface}: @tab @code{integer function omp_get_active_level()}
232 @item @emph{See also}:
233 @ref{omp_get_level}, @ref{omp_get_max_active_levels}, @ref{omp_set_max_active_levels}
235 @item @emph{Reference}:
236 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.20.
241 @node omp_get_ancestor_thread_num
242 @section @code{omp_get_ancestor_thread_num} -- Ancestor thread ID
244 @item @emph{Description}:
245 This function returns the thread identification number for the given
246 nesting level of the current thread. For values of @var{level} outside
247 zero to @code{omp_get_level} -1 is returned; if @var{level} is
248 @code{omp_get_level} the result is identical to @code{omp_get_thread_num}.
251 @multitable @columnfractions .20 .80
252 @item @emph{Prototype}: @tab @code{int omp_get_ancestor_thread_num(int level);}
255 @item @emph{Fortran}:
256 @multitable @columnfractions .20 .80
257 @item @emph{Interface}: @tab @code{integer function omp_get_ancestor_thread_num(level)}
258 @item @tab @code{integer level}
261 @item @emph{See also}:
262 @ref{omp_get_level}, @ref{omp_get_thread_num}, @ref{omp_get_team_size}
264 @item @emph{Reference}:
265 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.18.
270 @node omp_get_cancellation
271 @section @code{omp_get_cancellation} -- Whether cancellation support is enabled
273 @item @emph{Description}:
274 This function returns @code{true} if cancellation is activated, @code{false}
275 otherwise. Here, @code{true} and @code{false} represent their language-specific
276 counterparts. Unless @env{OMP_CANCELLATION} is set true, cancellations are
280 @multitable @columnfractions .20 .80
281 @item @emph{Prototype}: @tab @code{int omp_get_cancellation(void);}
284 @item @emph{Fortran}:
285 @multitable @columnfractions .20 .80
286 @item @emph{Interface}: @tab @code{logical function omp_get_cancellation()}
289 @item @emph{See also}:
290 @ref{OMP_CANCELLATION}
292 @item @emph{Reference}:
293 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.9.
298 @node omp_get_default_device
299 @section @code{omp_get_default_device} -- Get the default device for target regions
301 @item @emph{Description}:
302 Get the default device for target regions without device clause.
305 @multitable @columnfractions .20 .80
306 @item @emph{Prototype}: @tab @code{int omp_get_default_device(void);}
309 @item @emph{Fortran}:
310 @multitable @columnfractions .20 .80
311 @item @emph{Interface}: @tab @code{integer function omp_get_default_device()}
314 @item @emph{See also}:
315 @ref{OMP_DEFAULT_DEVICE}, @ref{omp_set_default_device}
317 @item @emph{Reference}:
318 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.30.
323 @node omp_get_dynamic
324 @section @code{omp_get_dynamic} -- Dynamic teams setting
326 @item @emph{Description}:
327 This function returns @code{true} if enabled, @code{false} otherwise.
328 Here, @code{true} and @code{false} represent their language-specific
331 The dynamic team setting may be initialized at startup by the
332 @env{OMP_DYNAMIC} environment variable or at runtime using
333 @code{omp_set_dynamic}. If undefined, dynamic adjustment is
337 @multitable @columnfractions .20 .80
338 @item @emph{Prototype}: @tab @code{int omp_get_dynamic(void);}
341 @item @emph{Fortran}:
342 @multitable @columnfractions .20 .80
343 @item @emph{Interface}: @tab @code{logical function omp_get_dynamic()}
346 @item @emph{See also}:
347 @ref{omp_set_dynamic}, @ref{OMP_DYNAMIC}
349 @item @emph{Reference}:
350 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.8.
356 @section @code{omp_get_level} -- Obtain the current nesting level
358 @item @emph{Description}:
359 This function returns the nesting level for the parallel blocks,
360 which enclose the calling call.
363 @multitable @columnfractions .20 .80
364 @item @emph{Prototype}: @tab @code{int omp_get_level(void);}
367 @item @emph{Fortran}:
368 @multitable @columnfractions .20 .80
369 @item @emph{Interface}: @tab @code{integer function omp_level()}
372 @item @emph{See also}:
373 @ref{omp_get_active_level}
375 @item @emph{Reference}:
376 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.17.
381 @node omp_get_max_active_levels
382 @section @code{omp_get_max_active_levels} -- Maximum number of active regions
384 @item @emph{Description}:
385 This function obtains the maximum allowed number of nested, active parallel regions.
388 @multitable @columnfractions .20 .80
389 @item @emph{Prototype}: @tab @code{int omp_get_max_active_levels(void);}
392 @item @emph{Fortran}:
393 @multitable @columnfractions .20 .80
394 @item @emph{Interface}: @tab @code{integer function omp_get_max_active_levels()}
397 @item @emph{See also}:
398 @ref{omp_set_max_active_levels}, @ref{omp_get_active_level}
400 @item @emph{Reference}:
401 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.16.
405 @node omp_get_max_task_priority
406 @section @code{omp_get_max_task_priority} -- Maximum priority value
407 that can be set for tasks.
409 @item @emph{Description}:
410 This function obtains the maximum allowed priority number for tasks.
413 @multitable @columnfractions .20 .80
414 @item @emph{Prototype}: @tab @code{int omp_get_max_task_priority(void);}
417 @item @emph{Fortran}:
418 @multitable @columnfractions .20 .80
419 @item @emph{Interface}: @tab @code{integer function omp_get_max_task_priority()}
422 @item @emph{Reference}:
423 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.29.
427 @node omp_get_max_threads
428 @section @code{omp_get_max_threads} -- Maximum number of threads of parallel region
430 @item @emph{Description}:
431 Return the maximum number of threads used for the current parallel region
432 that does not use the clause @code{num_threads}.
435 @multitable @columnfractions .20 .80
436 @item @emph{Prototype}: @tab @code{int omp_get_max_threads(void);}
439 @item @emph{Fortran}:
440 @multitable @columnfractions .20 .80
441 @item @emph{Interface}: @tab @code{integer function omp_get_max_threads()}
444 @item @emph{See also}:
445 @ref{omp_set_num_threads}, @ref{omp_set_dynamic}, @ref{omp_get_thread_limit}
447 @item @emph{Reference}:
448 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.3.
454 @section @code{omp_get_nested} -- Nested parallel regions
456 @item @emph{Description}:
457 This function returns @code{true} if nested parallel regions are
458 enabled, @code{false} otherwise. Here, @code{true} and @code{false}
459 represent their language-specific counterparts.
461 Nested parallel regions may be initialized at startup by the
462 @env{OMP_NESTED} environment variable or at runtime using
463 @code{omp_set_nested}. If undefined, nested parallel regions are
467 @multitable @columnfractions .20 .80
468 @item @emph{Prototype}: @tab @code{int omp_get_nested(void);}
471 @item @emph{Fortran}:
472 @multitable @columnfractions .20 .80
473 @item @emph{Interface}: @tab @code{logical function omp_get_nested()}
476 @item @emph{See also}:
477 @ref{omp_set_nested}, @ref{OMP_NESTED}
479 @item @emph{Reference}:
480 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.11.
485 @node omp_get_num_devices
486 @section @code{omp_get_num_devices} -- Number of target devices
488 @item @emph{Description}:
489 Returns the number of target devices.
492 @multitable @columnfractions .20 .80
493 @item @emph{Prototype}: @tab @code{int omp_get_num_devices(void);}
496 @item @emph{Fortran}:
497 @multitable @columnfractions .20 .80
498 @item @emph{Interface}: @tab @code{integer function omp_get_num_devices()}
501 @item @emph{Reference}:
502 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.31.
507 @node omp_get_num_procs
508 @section @code{omp_get_num_procs} -- Number of processors online
510 @item @emph{Description}:
511 Returns the number of processors online on that device.
514 @multitable @columnfractions .20 .80
515 @item @emph{Prototype}: @tab @code{int omp_get_num_procs(void);}
518 @item @emph{Fortran}:
519 @multitable @columnfractions .20 .80
520 @item @emph{Interface}: @tab @code{integer function omp_get_num_procs()}
523 @item @emph{Reference}:
524 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.5.
529 @node omp_get_num_teams
530 @section @code{omp_get_num_teams} -- Number of teams
532 @item @emph{Description}:
533 Returns the number of teams in the current team region.
536 @multitable @columnfractions .20 .80
537 @item @emph{Prototype}: @tab @code{int omp_get_num_teams(void);}
540 @item @emph{Fortran}:
541 @multitable @columnfractions .20 .80
542 @item @emph{Interface}: @tab @code{integer function omp_get_num_teams()}
545 @item @emph{Reference}:
546 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.32.
551 @node omp_get_num_threads
552 @section @code{omp_get_num_threads} -- Size of the active team
554 @item @emph{Description}:
555 Returns the number of threads in the current team. In a sequential section of
556 the program @code{omp_get_num_threads} returns 1.
558 The default team size may be initialized at startup by the
559 @env{OMP_NUM_THREADS} environment variable. At runtime, the size
560 of the current team may be set either by the @code{NUM_THREADS}
561 clause or by @code{omp_set_num_threads}. If none of the above were
562 used to define a specific value and @env{OMP_DYNAMIC} is disabled,
563 one thread per CPU online is used.
566 @multitable @columnfractions .20 .80
567 @item @emph{Prototype}: @tab @code{int omp_get_num_threads(void);}
570 @item @emph{Fortran}:
571 @multitable @columnfractions .20 .80
572 @item @emph{Interface}: @tab @code{integer function omp_get_num_threads()}
575 @item @emph{See also}:
576 @ref{omp_get_max_threads}, @ref{omp_set_num_threads}, @ref{OMP_NUM_THREADS}
578 @item @emph{Reference}:
579 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.2.
584 @node omp_get_proc_bind
585 @section @code{omp_get_proc_bind} -- Whether theads may be moved between CPUs
587 @item @emph{Description}:
588 This functions returns the currently active thread affinity policy, which is
589 set via @env{OMP_PROC_BIND}. Possible values are @code{omp_proc_bind_false},
590 @code{omp_proc_bind_true}, @code{omp_proc_bind_master},
591 @code{omp_proc_bind_close} and @code{omp_proc_bind_spread}.
594 @multitable @columnfractions .20 .80
595 @item @emph{Prototype}: @tab @code{omp_proc_bind_t omp_get_proc_bind(void);}
598 @item @emph{Fortran}:
599 @multitable @columnfractions .20 .80
600 @item @emph{Interface}: @tab @code{integer(kind=omp_proc_bind_kind) function omp_get_proc_bind()}
603 @item @emph{See also}:
604 @ref{OMP_PROC_BIND}, @ref{OMP_PLACES}, @ref{GOMP_CPU_AFFINITY},
606 @item @emph{Reference}:
607 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.22.
612 @node omp_get_schedule
613 @section @code{omp_get_schedule} -- Obtain the runtime scheduling method
615 @item @emph{Description}:
616 Obtain the runtime scheduling method. The @var{kind} argument will be
617 set to the value @code{omp_sched_static}, @code{omp_sched_dynamic},
618 @code{omp_sched_guided} or @code{omp_sched_auto}. The second argument,
619 @var{chunk_size}, is set to the chunk size.
622 @multitable @columnfractions .20 .80
623 @item @emph{Prototype}: @tab @code{void omp_get_schedule(omp_sched_t *kind, int *chunk_size);}
626 @item @emph{Fortran}:
627 @multitable @columnfractions .20 .80
628 @item @emph{Interface}: @tab @code{subroutine omp_get_schedule(kind, chunk_size)}
629 @item @tab @code{integer(kind=omp_sched_kind) kind}
630 @item @tab @code{integer chunk_size}
633 @item @emph{See also}:
634 @ref{omp_set_schedule}, @ref{OMP_SCHEDULE}
636 @item @emph{Reference}:
637 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.13.
642 @node omp_get_team_num
643 @section @code{omp_get_team_num} -- Get team number
645 @item @emph{Description}:
646 Returns the team number of the calling thread.
649 @multitable @columnfractions .20 .80
650 @item @emph{Prototype}: @tab @code{int omp_get_team_num(void);}
653 @item @emph{Fortran}:
654 @multitable @columnfractions .20 .80
655 @item @emph{Interface}: @tab @code{integer function omp_get_team_num()}
658 @item @emph{Reference}:
659 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.33.
664 @node omp_get_team_size
665 @section @code{omp_get_team_size} -- Number of threads in a team
667 @item @emph{Description}:
668 This function returns the number of threads in a thread team to which
669 either the current thread or its ancestor belongs. For values of @var{level}
670 outside zero to @code{omp_get_level}, -1 is returned; if @var{level} is zero,
671 1 is returned, and for @code{omp_get_level}, the result is identical
672 to @code{omp_get_num_threads}.
675 @multitable @columnfractions .20 .80
676 @item @emph{Prototype}: @tab @code{int omp_get_team_size(int level);}
679 @item @emph{Fortran}:
680 @multitable @columnfractions .20 .80
681 @item @emph{Interface}: @tab @code{integer function omp_get_team_size(level)}
682 @item @tab @code{integer level}
685 @item @emph{See also}:
686 @ref{omp_get_num_threads}, @ref{omp_get_level}, @ref{omp_get_ancestor_thread_num}
688 @item @emph{Reference}:
689 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.19.
694 @node omp_get_thread_limit
695 @section @code{omp_get_thread_limit} -- Maximum number of threads
697 @item @emph{Description}:
698 Return the maximum number of threads of the program.
701 @multitable @columnfractions .20 .80
702 @item @emph{Prototype}: @tab @code{int omp_get_thread_limit(void);}
705 @item @emph{Fortran}:
706 @multitable @columnfractions .20 .80
707 @item @emph{Interface}: @tab @code{integer function omp_get_thread_limit()}
710 @item @emph{See also}:
711 @ref{omp_get_max_threads}, @ref{OMP_THREAD_LIMIT}
713 @item @emph{Reference}:
714 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.14.
719 @node omp_get_thread_num
720 @section @code{omp_get_thread_num} -- Current thread ID
722 @item @emph{Description}:
723 Returns a unique thread identification number within the current team.
724 In a sequential parts of the program, @code{omp_get_thread_num}
725 always returns 0. In parallel regions the return value varies
726 from 0 to @code{omp_get_num_threads}-1 inclusive. The return
727 value of the master thread of a team is always 0.
730 @multitable @columnfractions .20 .80
731 @item @emph{Prototype}: @tab @code{int omp_get_thread_num(void);}
734 @item @emph{Fortran}:
735 @multitable @columnfractions .20 .80
736 @item @emph{Interface}: @tab @code{integer function omp_get_thread_num()}
739 @item @emph{See also}:
740 @ref{omp_get_num_threads}, @ref{omp_get_ancestor_thread_num}
742 @item @emph{Reference}:
743 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.4.
748 @node omp_in_parallel
749 @section @code{omp_in_parallel} -- Whether a parallel region is active
751 @item @emph{Description}:
752 This function returns @code{true} if currently running in parallel,
753 @code{false} otherwise. Here, @code{true} and @code{false} represent
754 their language-specific counterparts.
757 @multitable @columnfractions .20 .80
758 @item @emph{Prototype}: @tab @code{int omp_in_parallel(void);}
761 @item @emph{Fortran}:
762 @multitable @columnfractions .20 .80
763 @item @emph{Interface}: @tab @code{logical function omp_in_parallel()}
766 @item @emph{Reference}:
767 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.6.
772 @section @code{omp_in_final} -- Whether in final or included task region
774 @item @emph{Description}:
775 This function returns @code{true} if currently running in a final
776 or included task region, @code{false} otherwise. Here, @code{true}
777 and @code{false} represent their language-specific counterparts.
780 @multitable @columnfractions .20 .80
781 @item @emph{Prototype}: @tab @code{int omp_in_final(void);}
784 @item @emph{Fortran}:
785 @multitable @columnfractions .20 .80
786 @item @emph{Interface}: @tab @code{logical function omp_in_final()}
789 @item @emph{Reference}:
790 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.21.
795 @node omp_is_initial_device
796 @section @code{omp_is_initial_device} -- Whether executing on the host device
798 @item @emph{Description}:
799 This function returns @code{true} if currently running on the host device,
800 @code{false} otherwise. Here, @code{true} and @code{false} represent
801 their language-specific counterparts.
804 @multitable @columnfractions .20 .80
805 @item @emph{Prototype}: @tab @code{int omp_is_initial_device(void);}
808 @item @emph{Fortran}:
809 @multitable @columnfractions .20 .80
810 @item @emph{Interface}: @tab @code{logical function omp_is_initial_device()}
813 @item @emph{Reference}:
814 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.34.
819 @node omp_set_default_device
820 @section @code{omp_set_default_device} -- Set the default device for target regions
822 @item @emph{Description}:
823 Set the default device for target regions without device clause. The argument
824 shall be a nonnegative device number.
827 @multitable @columnfractions .20 .80
828 @item @emph{Prototype}: @tab @code{void omp_set_default_device(int device_num);}
831 @item @emph{Fortran}:
832 @multitable @columnfractions .20 .80
833 @item @emph{Interface}: @tab @code{subroutine omp_set_default_device(device_num)}
834 @item @tab @code{integer device_num}
837 @item @emph{See also}:
838 @ref{OMP_DEFAULT_DEVICE}, @ref{omp_get_default_device}
840 @item @emph{Reference}:
841 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.29.
846 @node omp_set_dynamic
847 @section @code{omp_set_dynamic} -- Enable/disable dynamic teams
849 @item @emph{Description}:
850 Enable or disable the dynamic adjustment of the number of threads
851 within a team. The function takes the language-specific equivalent
852 of @code{true} and @code{false}, where @code{true} enables dynamic
853 adjustment of team sizes and @code{false} disables it.
856 @multitable @columnfractions .20 .80
857 @item @emph{Prototype}: @tab @code{void omp_set_dynamic(int dynamic_threads);}
860 @item @emph{Fortran}:
861 @multitable @columnfractions .20 .80
862 @item @emph{Interface}: @tab @code{subroutine omp_set_dynamic(dynamic_threads)}
863 @item @tab @code{logical, intent(in) :: dynamic_threads}
866 @item @emph{See also}:
867 @ref{OMP_DYNAMIC}, @ref{omp_get_dynamic}
869 @item @emph{Reference}:
870 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.7.
875 @node omp_set_max_active_levels
876 @section @code{omp_set_max_active_levels} -- Limits the number of active parallel regions
878 @item @emph{Description}:
879 This function limits the maximum allowed number of nested, active
883 @multitable @columnfractions .20 .80
884 @item @emph{Prototype}: @tab @code{void omp_set_max_active_levels(int max_levels);}
887 @item @emph{Fortran}:
888 @multitable @columnfractions .20 .80
889 @item @emph{Interface}: @tab @code{subroutine omp_set_max_active_levels(max_levels)}
890 @item @tab @code{integer max_levels}
893 @item @emph{See also}:
894 @ref{omp_get_max_active_levels}, @ref{omp_get_active_level}
896 @item @emph{Reference}:
897 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.15.
903 @section @code{omp_set_nested} -- Enable/disable nested parallel regions
905 @item @emph{Description}:
906 Enable or disable nested parallel regions, i.e., whether team members
907 are allowed to create new teams. The function takes the language-specific
908 equivalent of @code{true} and @code{false}, where @code{true} enables
909 dynamic adjustment of team sizes and @code{false} disables it.
912 @multitable @columnfractions .20 .80
913 @item @emph{Prototype}: @tab @code{void omp_set_nested(int nested);}
916 @item @emph{Fortran}:
917 @multitable @columnfractions .20 .80
918 @item @emph{Interface}: @tab @code{subroutine omp_set_nested(nested)}
919 @item @tab @code{logical, intent(in) :: nested}
922 @item @emph{See also}:
923 @ref{OMP_NESTED}, @ref{omp_get_nested}
925 @item @emph{Reference}:
926 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.10.
931 @node omp_set_num_threads
932 @section @code{omp_set_num_threads} -- Set upper team size limit
934 @item @emph{Description}:
935 Specifies the number of threads used by default in subsequent parallel
936 sections, if those do not specify a @code{num_threads} clause. The
937 argument of @code{omp_set_num_threads} shall be a positive integer.
940 @multitable @columnfractions .20 .80
941 @item @emph{Prototype}: @tab @code{void omp_set_num_threads(int num_threads);}
944 @item @emph{Fortran}:
945 @multitable @columnfractions .20 .80
946 @item @emph{Interface}: @tab @code{subroutine omp_set_num_threads(num_threads)}
947 @item @tab @code{integer, intent(in) :: num_threads}
950 @item @emph{See also}:
951 @ref{OMP_NUM_THREADS}, @ref{omp_get_num_threads}, @ref{omp_get_max_threads}
953 @item @emph{Reference}:
954 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.1.
959 @node omp_set_schedule
960 @section @code{omp_set_schedule} -- Set the runtime scheduling method
962 @item @emph{Description}:
963 Sets the runtime scheduling method. The @var{kind} argument can have the
964 value @code{omp_sched_static}, @code{omp_sched_dynamic},
965 @code{omp_sched_guided} or @code{omp_sched_auto}. Except for
966 @code{omp_sched_auto}, the chunk size is set to the value of
967 @var{chunk_size} if positive, or to the default value if zero or negative.
968 For @code{omp_sched_auto} the @var{chunk_size} argument is ignored.
971 @multitable @columnfractions .20 .80
972 @item @emph{Prototype}: @tab @code{void omp_set_schedule(omp_sched_t kind, int chunk_size);}
975 @item @emph{Fortran}:
976 @multitable @columnfractions .20 .80
977 @item @emph{Interface}: @tab @code{subroutine omp_set_schedule(kind, chunk_size)}
978 @item @tab @code{integer(kind=omp_sched_kind) kind}
979 @item @tab @code{integer chunk_size}
982 @item @emph{See also}:
983 @ref{omp_get_schedule}
986 @item @emph{Reference}:
987 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.2.12.
993 @section @code{omp_init_lock} -- Initialize simple lock
995 @item @emph{Description}:
996 Initialize a simple lock. After initialization, the lock is in
1000 @multitable @columnfractions .20 .80
1001 @item @emph{Prototype}: @tab @code{void omp_init_lock(omp_lock_t *lock);}
1004 @item @emph{Fortran}:
1005 @multitable @columnfractions .20 .80
1006 @item @emph{Interface}: @tab @code{subroutine omp_init_lock(svar)}
1007 @item @tab @code{integer(omp_lock_kind), intent(out) :: svar}
1010 @item @emph{See also}:
1011 @ref{omp_destroy_lock}
1013 @item @emph{Reference}:
1014 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.1.
1020 @section @code{omp_set_lock} -- Wait for and set simple lock
1022 @item @emph{Description}:
1023 Before setting a simple lock, the lock variable must be initialized by
1024 @code{omp_init_lock}. The calling thread is blocked until the lock
1025 is available. If the lock is already held by the current thread,
1029 @multitable @columnfractions .20 .80
1030 @item @emph{Prototype}: @tab @code{void omp_set_lock(omp_lock_t *lock);}
1033 @item @emph{Fortran}:
1034 @multitable @columnfractions .20 .80
1035 @item @emph{Interface}: @tab @code{subroutine omp_set_lock(svar)}
1036 @item @tab @code{integer(omp_lock_kind), intent(inout) :: svar}
1039 @item @emph{See also}:
1040 @ref{omp_init_lock}, @ref{omp_test_lock}, @ref{omp_unset_lock}
1042 @item @emph{Reference}:
1043 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.4.
1049 @section @code{omp_test_lock} -- Test and set simple lock if available
1051 @item @emph{Description}:
1052 Before setting a simple lock, the lock variable must be initialized by
1053 @code{omp_init_lock}. Contrary to @code{omp_set_lock}, @code{omp_test_lock}
1054 does not block if the lock is not available. This function returns
1055 @code{true} upon success, @code{false} otherwise. Here, @code{true} and
1056 @code{false} represent their language-specific counterparts.
1059 @multitable @columnfractions .20 .80
1060 @item @emph{Prototype}: @tab @code{int omp_test_lock(omp_lock_t *lock);}
1063 @item @emph{Fortran}:
1064 @multitable @columnfractions .20 .80
1065 @item @emph{Interface}: @tab @code{logical function omp_test_lock(svar)}
1066 @item @tab @code{integer(omp_lock_kind), intent(inout) :: svar}
1069 @item @emph{See also}:
1070 @ref{omp_init_lock}, @ref{omp_set_lock}, @ref{omp_set_lock}
1072 @item @emph{Reference}:
1073 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.6.
1078 @node omp_unset_lock
1079 @section @code{omp_unset_lock} -- Unset simple lock
1081 @item @emph{Description}:
1082 A simple lock about to be unset must have been locked by @code{omp_set_lock}
1083 or @code{omp_test_lock} before. In addition, the lock must be held by the
1084 thread calling @code{omp_unset_lock}. Then, the lock becomes unlocked. If one
1085 or more threads attempted to set the lock before, one of them is chosen to,
1086 again, set the lock to itself.
1089 @multitable @columnfractions .20 .80
1090 @item @emph{Prototype}: @tab @code{void omp_unset_lock(omp_lock_t *lock);}
1093 @item @emph{Fortran}:
1094 @multitable @columnfractions .20 .80
1095 @item @emph{Interface}: @tab @code{subroutine omp_unset_lock(svar)}
1096 @item @tab @code{integer(omp_lock_kind), intent(inout) :: svar}
1099 @item @emph{See also}:
1100 @ref{omp_set_lock}, @ref{omp_test_lock}
1102 @item @emph{Reference}:
1103 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.5.
1108 @node omp_destroy_lock
1109 @section @code{omp_destroy_lock} -- Destroy simple lock
1111 @item @emph{Description}:
1112 Destroy a simple lock. In order to be destroyed, a simple lock must be
1113 in the unlocked state.
1116 @multitable @columnfractions .20 .80
1117 @item @emph{Prototype}: @tab @code{void omp_destroy_lock(omp_lock_t *lock);}
1120 @item @emph{Fortran}:
1121 @multitable @columnfractions .20 .80
1122 @item @emph{Interface}: @tab @code{subroutine omp_destroy_lock(svar)}
1123 @item @tab @code{integer(omp_lock_kind), intent(inout) :: svar}
1126 @item @emph{See also}:
1129 @item @emph{Reference}:
1130 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.3.
1135 @node omp_init_nest_lock
1136 @section @code{omp_init_nest_lock} -- Initialize nested lock
1138 @item @emph{Description}:
1139 Initialize a nested lock. After initialization, the lock is in
1140 an unlocked state and the nesting count is set to zero.
1143 @multitable @columnfractions .20 .80
1144 @item @emph{Prototype}: @tab @code{void omp_init_nest_lock(omp_nest_lock_t *lock);}
1147 @item @emph{Fortran}:
1148 @multitable @columnfractions .20 .80
1149 @item @emph{Interface}: @tab @code{subroutine omp_init_nest_lock(nvar)}
1150 @item @tab @code{integer(omp_nest_lock_kind), intent(out) :: nvar}
1153 @item @emph{See also}:
1154 @ref{omp_destroy_nest_lock}
1156 @item @emph{Reference}:
1157 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.1.
1161 @node omp_set_nest_lock
1162 @section @code{omp_set_nest_lock} -- Wait for and set nested lock
1164 @item @emph{Description}:
1165 Before setting a nested lock, the lock variable must be initialized by
1166 @code{omp_init_nest_lock}. The calling thread is blocked until the lock
1167 is available. If the lock is already held by the current thread, the
1168 nesting count for the lock is incremented.
1171 @multitable @columnfractions .20 .80
1172 @item @emph{Prototype}: @tab @code{void omp_set_nest_lock(omp_nest_lock_t *lock);}
1175 @item @emph{Fortran}:
1176 @multitable @columnfractions .20 .80
1177 @item @emph{Interface}: @tab @code{subroutine omp_set_nest_lock(nvar)}
1178 @item @tab @code{integer(omp_nest_lock_kind), intent(inout) :: nvar}
1181 @item @emph{See also}:
1182 @ref{omp_init_nest_lock}, @ref{omp_unset_nest_lock}
1184 @item @emph{Reference}:
1185 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.4.
1190 @node omp_test_nest_lock
1191 @section @code{omp_test_nest_lock} -- Test and set nested lock if available
1193 @item @emph{Description}:
1194 Before setting a nested lock, the lock variable must be initialized by
1195 @code{omp_init_nest_lock}. Contrary to @code{omp_set_nest_lock},
1196 @code{omp_test_nest_lock} does not block if the lock is not available.
1197 If the lock is already held by the current thread, the new nesting count
1198 is returned. Otherwise, the return value equals zero.
1201 @multitable @columnfractions .20 .80
1202 @item @emph{Prototype}: @tab @code{int omp_test_nest_lock(omp_nest_lock_t *lock);}
1205 @item @emph{Fortran}:
1206 @multitable @columnfractions .20 .80
1207 @item @emph{Interface}: @tab @code{logical function omp_test_nest_lock(nvar)}
1208 @item @tab @code{integer(omp_nest_lock_kind), intent(inout) :: nvar}
1212 @item @emph{See also}:
1213 @ref{omp_init_lock}, @ref{omp_set_lock}, @ref{omp_set_lock}
1215 @item @emph{Reference}:
1216 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.6.
1221 @node omp_unset_nest_lock
1222 @section @code{omp_unset_nest_lock} -- Unset nested lock
1224 @item @emph{Description}:
1225 A nested lock about to be unset must have been locked by @code{omp_set_nested_lock}
1226 or @code{omp_test_nested_lock} before. In addition, the lock must be held by the
1227 thread calling @code{omp_unset_nested_lock}. If the nesting count drops to zero, the
1228 lock becomes unlocked. If one ore more threads attempted to set the lock before,
1229 one of them is chosen to, again, set the lock to itself.
1232 @multitable @columnfractions .20 .80
1233 @item @emph{Prototype}: @tab @code{void omp_unset_nest_lock(omp_nest_lock_t *lock);}
1236 @item @emph{Fortran}:
1237 @multitable @columnfractions .20 .80
1238 @item @emph{Interface}: @tab @code{subroutine omp_unset_nest_lock(nvar)}
1239 @item @tab @code{integer(omp_nest_lock_kind), intent(inout) :: nvar}
1242 @item @emph{See also}:
1243 @ref{omp_set_nest_lock}
1245 @item @emph{Reference}:
1246 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.5.
1251 @node omp_destroy_nest_lock
1252 @section @code{omp_destroy_nest_lock} -- Destroy nested lock
1254 @item @emph{Description}:
1255 Destroy a nested lock. In order to be destroyed, a nested lock must be
1256 in the unlocked state and its nesting count must equal zero.
1259 @multitable @columnfractions .20 .80
1260 @item @emph{Prototype}: @tab @code{void omp_destroy_nest_lock(omp_nest_lock_t *);}
1263 @item @emph{Fortran}:
1264 @multitable @columnfractions .20 .80
1265 @item @emph{Interface}: @tab @code{subroutine omp_destroy_nest_lock(nvar)}
1266 @item @tab @code{integer(omp_nest_lock_kind), intent(inout) :: nvar}
1269 @item @emph{See also}:
1272 @item @emph{Reference}:
1273 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.3.3.
1279 @section @code{omp_get_wtick} -- Get timer precision
1281 @item @emph{Description}:
1282 Gets the timer precision, i.e., the number of seconds between two
1283 successive clock ticks.
1286 @multitable @columnfractions .20 .80
1287 @item @emph{Prototype}: @tab @code{double omp_get_wtick(void);}
1290 @item @emph{Fortran}:
1291 @multitable @columnfractions .20 .80
1292 @item @emph{Interface}: @tab @code{double precision function omp_get_wtick()}
1295 @item @emph{See also}:
1298 @item @emph{Reference}:
1299 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.4.2.
1305 @section @code{omp_get_wtime} -- Elapsed wall clock time
1307 @item @emph{Description}:
1308 Elapsed wall clock time in seconds. The time is measured per thread, no
1309 guarantee can be made that two distinct threads measure the same time.
1310 Time is measured from some "time in the past", which is an arbitrary time
1311 guaranteed not to change during the execution of the program.
1314 @multitable @columnfractions .20 .80
1315 @item @emph{Prototype}: @tab @code{double omp_get_wtime(void);}
1318 @item @emph{Fortran}:
1319 @multitable @columnfractions .20 .80
1320 @item @emph{Interface}: @tab @code{double precision function omp_get_wtime()}
1323 @item @emph{See also}:
1326 @item @emph{Reference}:
1327 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 3.4.1.
1332 @c ---------------------------------------------------------------------
1333 @c OpenMP Environment Variables
1334 @c ---------------------------------------------------------------------
1336 @node Environment Variables
1337 @chapter OpenMP Environment Variables
1339 The environment variables which beginning with @env{OMP_} are defined by
1340 section 4 of the OpenMP specification in version 4.5, while those
1341 beginning with @env{GOMP_} are GNU extensions.
1344 * OMP_CANCELLATION:: Set whether cancellation is activated
1345 * OMP_DISPLAY_ENV:: Show OpenMP version and environment variables
1346 * OMP_DEFAULT_DEVICE:: Set the device used in target regions
1347 * OMP_DYNAMIC:: Dynamic adjustment of threads
1348 * OMP_MAX_ACTIVE_LEVELS:: Set the maximum number of nested parallel regions
1349 * OMP_MAX_TASK_PRIORITY:: Set the maximum task priority value
1350 * OMP_NESTED:: Nested parallel regions
1351 * OMP_NUM_THREADS:: Specifies the number of threads to use
1352 * OMP_PROC_BIND:: Whether theads may be moved between CPUs
1353 * OMP_PLACES:: Specifies on which CPUs the theads should be placed
1354 * OMP_STACKSIZE:: Set default thread stack size
1355 * OMP_SCHEDULE:: How threads are scheduled
1356 * OMP_THREAD_LIMIT:: Set the maximum number of threads
1357 * OMP_WAIT_POLICY:: How waiting threads are handled
1358 * GOMP_CPU_AFFINITY:: Bind threads to specific CPUs
1359 * GOMP_DEBUG:: Enable debugging output
1360 * GOMP_STACKSIZE:: Set default thread stack size
1361 * GOMP_SPINCOUNT:: Set the busy-wait spin count
1362 * GOMP_RTEMS_THREAD_POOLS:: Set the RTEMS specific thread pools
1366 @node OMP_CANCELLATION
1367 @section @env{OMP_CANCELLATION} -- Set whether cancellation is activated
1368 @cindex Environment Variable
1370 @item @emph{Description}:
1371 If set to @code{TRUE}, the cancellation is activated. If set to @code{FALSE} or
1372 if unset, cancellation is disabled and the @code{cancel} construct is ignored.
1374 @item @emph{See also}:
1375 @ref{omp_get_cancellation}
1377 @item @emph{Reference}:
1378 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.11
1383 @node OMP_DISPLAY_ENV
1384 @section @env{OMP_DISPLAY_ENV} -- Show OpenMP version and environment variables
1385 @cindex Environment Variable
1387 @item @emph{Description}:
1388 If set to @code{TRUE}, the OpenMP version number and the values
1389 associated with the OpenMP environment variables are printed to @code{stderr}.
1390 If set to @code{VERBOSE}, it additionally shows the value of the environment
1391 variables which are GNU extensions. If undefined or set to @code{FALSE},
1392 this information will not be shown.
1395 @item @emph{Reference}:
1396 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.12
1401 @node OMP_DEFAULT_DEVICE
1402 @section @env{OMP_DEFAULT_DEVICE} -- Set the device used in target regions
1403 @cindex Environment Variable
1405 @item @emph{Description}:
1406 Set to choose the device which is used in a @code{target} region, unless the
1407 value is overridden by @code{omp_set_default_device} or by a @code{device}
1408 clause. The value shall be the nonnegative device number. If no device with
1409 the given device number exists, the code is executed on the host. If unset,
1410 device number 0 will be used.
1413 @item @emph{See also}:
1414 @ref{omp_get_default_device}, @ref{omp_set_default_device},
1416 @item @emph{Reference}:
1417 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.13
1423 @section @env{OMP_DYNAMIC} -- Dynamic adjustment of threads
1424 @cindex Environment Variable
1426 @item @emph{Description}:
1427 Enable or disable the dynamic adjustment of the number of threads
1428 within a team. The value of this environment variable shall be
1429 @code{TRUE} or @code{FALSE}. If undefined, dynamic adjustment is
1430 disabled by default.
1432 @item @emph{See also}:
1433 @ref{omp_set_dynamic}
1435 @item @emph{Reference}:
1436 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.3
1441 @node OMP_MAX_ACTIVE_LEVELS
1442 @section @env{OMP_MAX_ACTIVE_LEVELS} -- Set the maximum number of nested parallel regions
1443 @cindex Environment Variable
1445 @item @emph{Description}:
1446 Specifies the initial value for the maximum number of nested parallel
1447 regions. The value of this variable shall be a positive integer.
1448 If undefined, the number of active levels is unlimited.
1450 @item @emph{See also}:
1451 @ref{omp_set_max_active_levels}
1453 @item @emph{Reference}:
1454 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.9
1459 @node OMP_MAX_TASK_PRIORITY
1460 @section @env{OMP_MAX_TASK_PRIORITY} -- Set the maximum priority
1461 number that can be set for a task.
1462 @cindex Environment Variable
1464 @item @emph{Description}:
1465 Specifies the initial value for the maximum priority value that can be
1466 set for a task. The value of this variable shall be a non-negative
1467 integer, and zero is allowed. If undefined, the default priority is
1470 @item @emph{See also}:
1471 @ref{omp_get_max_task_priority}
1473 @item @emph{Reference}:
1474 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.14
1480 @section @env{OMP_NESTED} -- Nested parallel regions
1481 @cindex Environment Variable
1482 @cindex Implementation specific setting
1484 @item @emph{Description}:
1485 Enable or disable nested parallel regions, i.e., whether team members
1486 are allowed to create new teams. The value of this environment variable
1487 shall be @code{TRUE} or @code{FALSE}. If undefined, nested parallel
1488 regions are disabled by default.
1490 @item @emph{See also}:
1491 @ref{omp_set_nested}
1493 @item @emph{Reference}:
1494 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.6
1499 @node OMP_NUM_THREADS
1500 @section @env{OMP_NUM_THREADS} -- Specifies the number of threads to use
1501 @cindex Environment Variable
1502 @cindex Implementation specific setting
1504 @item @emph{Description}:
1505 Specifies the default number of threads to use in parallel regions. The
1506 value of this variable shall be a comma-separated list of positive integers;
1507 the value specified the number of threads to use for the corresponding nested
1508 level. If undefined one thread per CPU is used.
1510 @item @emph{See also}:
1511 @ref{omp_set_num_threads}
1513 @item @emph{Reference}:
1514 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.2
1520 @section @env{OMP_PROC_BIND} -- Whether theads may be moved between CPUs
1521 @cindex Environment Variable
1523 @item @emph{Description}:
1524 Specifies whether threads may be moved between processors. If set to
1525 @code{TRUE}, OpenMP theads should not be moved; if set to @code{FALSE}
1526 they may be moved. Alternatively, a comma separated list with the
1527 values @code{MASTER}, @code{CLOSE} and @code{SPREAD} can be used to specify
1528 the thread affinity policy for the corresponding nesting level. With
1529 @code{MASTER} the worker threads are in the same place partition as the
1530 master thread. With @code{CLOSE} those are kept close to the master thread
1531 in contiguous place partitions. And with @code{SPREAD} a sparse distribution
1532 across the place partitions is used.
1534 When undefined, @env{OMP_PROC_BIND} defaults to @code{TRUE} when
1535 @env{OMP_PLACES} or @env{GOMP_CPU_AFFINITY} is set and @code{FALSE} otherwise.
1537 @item @emph{See also}:
1538 @ref{OMP_PLACES}, @ref{GOMP_CPU_AFFINITY}, @ref{omp_get_proc_bind}
1540 @item @emph{Reference}:
1541 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.4
1547 @section @env{OMP_PLACES} -- Specifies on which CPUs the theads should be placed
1548 @cindex Environment Variable
1550 @item @emph{Description}:
1551 The thread placement can be either specified using an abstract name or by an
1552 explicit list of the places. The abstract names @code{threads}, @code{cores}
1553 and @code{sockets} can be optionally followed by a positive number in
1554 parentheses, which denotes the how many places shall be created. With
1555 @code{threads} each place corresponds to a single hardware thread; @code{cores}
1556 to a single core with the corresponding number of hardware threads; and with
1557 @code{sockets} the place corresponds to a single socket. The resulting
1558 placement can be shown by setting the @env{OMP_DISPLAY_ENV} environment
1561 Alternatively, the placement can be specified explicitly as comma-separated
1562 list of places. A place is specified by set of nonnegative numbers in curly
1563 braces, denoting the denoting the hardware threads. The hardware threads
1564 belonging to a place can either be specified as comma-separated list of
1565 nonnegative thread numbers or using an interval. Multiple places can also be
1566 either specified by a comma-separated list of places or by an interval. To
1567 specify an interval, a colon followed by the count is placed after after
1568 the hardware thread number or the place. Optionally, the length can be
1569 followed by a colon and the stride number -- otherwise a unit stride is
1570 assumed. For instance, the following specifies the same places list:
1571 @code{"@{0,1,2@}, @{3,4,6@}, @{7,8,9@}, @{10,11,12@}"};
1572 @code{"@{0:3@}, @{3:3@}, @{7:3@}, @{10:3@}"}; and @code{"@{0:2@}:4:3"}.
1574 If @env{OMP_PLACES} and @env{GOMP_CPU_AFFINITY} are unset and
1575 @env{OMP_PROC_BIND} is either unset or @code{false}, threads may be moved
1576 between CPUs following no placement policy.
1578 @item @emph{See also}:
1579 @ref{OMP_PROC_BIND}, @ref{GOMP_CPU_AFFINITY}, @ref{omp_get_proc_bind},
1580 @ref{OMP_DISPLAY_ENV}
1582 @item @emph{Reference}:
1583 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.5
1589 @section @env{OMP_STACKSIZE} -- Set default thread stack size
1590 @cindex Environment Variable
1592 @item @emph{Description}:
1593 Set the default thread stack size in kilobytes, unless the number
1594 is suffixed by @code{B}, @code{K}, @code{M} or @code{G}, in which
1595 case the size is, respectively, in bytes, kilobytes, megabytes
1596 or gigabytes. This is different from @code{pthread_attr_setstacksize}
1597 which gets the number of bytes as an argument. If the stack size cannot
1598 be set due to system constraints, an error is reported and the initial
1599 stack size is left unchanged. If undefined, the stack size is system
1602 @item @emph{Reference}:
1603 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.7
1609 @section @env{OMP_SCHEDULE} -- How threads are scheduled
1610 @cindex Environment Variable
1611 @cindex Implementation specific setting
1613 @item @emph{Description}:
1614 Allows to specify @code{schedule type} and @code{chunk size}.
1615 The value of the variable shall have the form: @code{type[,chunk]} where
1616 @code{type} is one of @code{static}, @code{dynamic}, @code{guided} or @code{auto}
1617 The optional @code{chunk} size shall be a positive integer. If undefined,
1618 dynamic scheduling and a chunk size of 1 is used.
1620 @item @emph{See also}:
1621 @ref{omp_set_schedule}
1623 @item @emph{Reference}:
1624 @uref{https://www.openmp.org, OpenMP specification v4.5}, Sections 2.7.1.1 and 4.1
1629 @node OMP_THREAD_LIMIT
1630 @section @env{OMP_THREAD_LIMIT} -- Set the maximum number of threads
1631 @cindex Environment Variable
1633 @item @emph{Description}:
1634 Specifies the number of threads to use for the whole program. The
1635 value of this variable shall be a positive integer. If undefined,
1636 the number of threads is not limited.
1638 @item @emph{See also}:
1639 @ref{OMP_NUM_THREADS}, @ref{omp_get_thread_limit}
1641 @item @emph{Reference}:
1642 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.10
1647 @node OMP_WAIT_POLICY
1648 @section @env{OMP_WAIT_POLICY} -- How waiting threads are handled
1649 @cindex Environment Variable
1651 @item @emph{Description}:
1652 Specifies whether waiting threads should be active or passive. If
1653 the value is @code{PASSIVE}, waiting threads should not consume CPU
1654 power while waiting; while the value is @code{ACTIVE} specifies that
1655 they should. If undefined, threads wait actively for a short time
1656 before waiting passively.
1658 @item @emph{See also}:
1659 @ref{GOMP_SPINCOUNT}
1661 @item @emph{Reference}:
1662 @uref{https://www.openmp.org, OpenMP specification v4.5}, Section 4.8
1667 @node GOMP_CPU_AFFINITY
1668 @section @env{GOMP_CPU_AFFINITY} -- Bind threads to specific CPUs
1669 @cindex Environment Variable
1671 @item @emph{Description}:
1672 Binds threads to specific CPUs. The variable should contain a space-separated
1673 or comma-separated list of CPUs. This list may contain different kinds of
1674 entries: either single CPU numbers in any order, a range of CPUs (M-N)
1675 or a range with some stride (M-N:S). CPU numbers are zero based. For example,
1676 @code{GOMP_CPU_AFFINITY="0 3 1-2 4-15:2"} will bind the initial thread
1677 to CPU 0, the second to CPU 3, the third to CPU 1, the fourth to
1678 CPU 2, the fifth to CPU 4, the sixth through tenth to CPUs 6, 8, 10, 12,
1679 and 14 respectively and then start assigning back from the beginning of
1680 the list. @code{GOMP_CPU_AFFINITY=0} binds all threads to CPU 0.
1682 There is no libgomp library routine to determine whether a CPU affinity
1683 specification is in effect. As a workaround, language-specific library
1684 functions, e.g., @code{getenv} in C or @code{GET_ENVIRONMENT_VARIABLE} in
1685 Fortran, may be used to query the setting of the @code{GOMP_CPU_AFFINITY}
1686 environment variable. A defined CPU affinity on startup cannot be changed
1687 or disabled during the runtime of the application.
1689 If both @env{GOMP_CPU_AFFINITY} and @env{OMP_PROC_BIND} are set,
1690 @env{OMP_PROC_BIND} has a higher precedence. If neither has been set and
1691 @env{OMP_PROC_BIND} is unset, or when @env{OMP_PROC_BIND} is set to
1692 @code{FALSE}, the host system will handle the assignment of threads to CPUs.
1694 @item @emph{See also}:
1695 @ref{OMP_PLACES}, @ref{OMP_PROC_BIND}
1701 @section @env{GOMP_DEBUG} -- Enable debugging output
1702 @cindex Environment Variable
1704 @item @emph{Description}:
1705 Enable debugging output. The variable should be set to @code{0}
1706 (disabled, also the default if not set), or @code{1} (enabled).
1708 If enabled, some debugging output will be printed during execution.
1709 This is currently not specified in more detail, and subject to change.
1714 @node GOMP_STACKSIZE
1715 @section @env{GOMP_STACKSIZE} -- Set default thread stack size
1716 @cindex Environment Variable
1717 @cindex Implementation specific setting
1719 @item @emph{Description}:
1720 Set the default thread stack size in kilobytes. This is different from
1721 @code{pthread_attr_setstacksize} which gets the number of bytes as an
1722 argument. If the stack size cannot be set due to system constraints, an
1723 error is reported and the initial stack size is left unchanged. If undefined,
1724 the stack size is system dependent.
1726 @item @emph{See also}:
1729 @item @emph{Reference}:
1730 @uref{https://gcc.gnu.org/ml/gcc-patches/2006-06/msg00493.html,
1731 GCC Patches Mailinglist},
1732 @uref{https://gcc.gnu.org/ml/gcc-patches/2006-06/msg00496.html,
1733 GCC Patches Mailinglist}
1738 @node GOMP_SPINCOUNT
1739 @section @env{GOMP_SPINCOUNT} -- Set the busy-wait spin count
1740 @cindex Environment Variable
1741 @cindex Implementation specific setting
1743 @item @emph{Description}:
1744 Determines how long a threads waits actively with consuming CPU power
1745 before waiting passively without consuming CPU power. The value may be
1746 either @code{INFINITE}, @code{INFINITY} to always wait actively or an
1747 integer which gives the number of spins of the busy-wait loop. The
1748 integer may optionally be followed by the following suffixes acting
1749 as multiplication factors: @code{k} (kilo, thousand), @code{M} (mega,
1750 million), @code{G} (giga, billion), or @code{T} (tera, trillion).
1751 If undefined, 0 is used when @env{OMP_WAIT_POLICY} is @code{PASSIVE},
1752 300,000 is used when @env{OMP_WAIT_POLICY} is undefined and
1753 30 billion is used when @env{OMP_WAIT_POLICY} is @code{ACTIVE}.
1754 If there are more OpenMP threads than available CPUs, 1000 and 100
1755 spins are used for @env{OMP_WAIT_POLICY} being @code{ACTIVE} or
1756 undefined, respectively; unless the @env{GOMP_SPINCOUNT} is lower
1757 or @env{OMP_WAIT_POLICY} is @code{PASSIVE}.
1759 @item @emph{See also}:
1760 @ref{OMP_WAIT_POLICY}
1765 @node GOMP_RTEMS_THREAD_POOLS
1766 @section @env{GOMP_RTEMS_THREAD_POOLS} -- Set the RTEMS specific thread pools
1767 @cindex Environment Variable
1768 @cindex Implementation specific setting
1770 @item @emph{Description}:
1771 This environment variable is only used on the RTEMS real-time operating system.
1772 It determines the scheduler instance specific thread pools. The format for
1773 @env{GOMP_RTEMS_THREAD_POOLS} is a list of optional
1774 @code{<thread-pool-count>[$<priority>]@@<scheduler-name>} configurations
1775 separated by @code{:} where:
1777 @item @code{<thread-pool-count>} is the thread pool count for this scheduler
1779 @item @code{$<priority>} is an optional priority for the worker threads of a
1780 thread pool according to @code{pthread_setschedparam}. In case a priority
1781 value is omitted, then a worker thread will inherit the priority of the OpenMP
1782 master thread that created it. The priority of the worker thread is not
1783 changed after creation, even if a new OpenMP master thread using the worker has
1784 a different priority.
1785 @item @code{@@<scheduler-name>} is the scheduler instance name according to the
1786 RTEMS application configuration.
1788 In case no thread pool configuration is specified for a scheduler instance,
1789 then each OpenMP master thread of this scheduler instance will use its own
1790 dynamically allocated thread pool. To limit the worker thread count of the
1791 thread pools, each OpenMP master thread must call @code{omp_set_num_threads}.
1792 @item @emph{Example}:
1793 Lets suppose we have three scheduler instances @code{IO}, @code{WRK0}, and
1794 @code{WRK1} with @env{GOMP_RTEMS_THREAD_POOLS} set to
1795 @code{"1@@WRK0:3$4@@WRK1"}. Then there are no thread pool restrictions for
1796 scheduler instance @code{IO}. In the scheduler instance @code{WRK0} there is
1797 one thread pool available. Since no priority is specified for this scheduler
1798 instance, the worker thread inherits the priority of the OpenMP master thread
1799 that created it. In the scheduler instance @code{WRK1} there are three thread
1800 pools available and their worker threads run at priority four.
1805 @c ---------------------------------------------------------------------
1807 @c ---------------------------------------------------------------------
1809 @node Enabling OpenACC
1810 @chapter Enabling OpenACC
1812 To activate the OpenACC extensions for C/C++ and Fortran, the compile-time
1813 flag @option{-fopenacc} must be specified. This enables the OpenACC directive
1814 @code{#pragma acc} in C/C++ and @code{!$acc} directives in free form,
1815 @code{c$acc}, @code{*$acc} and @code{!$acc} directives in fixed form,
1816 @code{!$} conditional compilation sentinels in free form and @code{c$},
1817 @code{*$} and @code{!$} sentinels in fixed form, for Fortran. The flag also
1818 arranges for automatic linking of the OpenACC runtime library
1819 (@ref{OpenACC Runtime Library Routines}).
1821 See @uref{https://gcc.gnu.org/wiki/OpenACC} for more information.
1823 A complete description of all OpenACC directives accepted may be found in
1824 the @uref{https://www.openacc.org, OpenACC} Application Programming
1825 Interface manual, version 2.6.
1829 @c ---------------------------------------------------------------------
1830 @c OpenACC Runtime Library Routines
1831 @c ---------------------------------------------------------------------
1833 @node OpenACC Runtime Library Routines
1834 @chapter OpenACC Runtime Library Routines
1836 The runtime routines described here are defined by section 3 of the OpenACC
1837 specifications in version 2.6.
1838 They have C linkage, and do not throw exceptions.
1839 Generally, they are available only for the host, with the exception of
1840 @code{acc_on_device}, which is available for both the host and the
1841 acceleration device.
1844 * acc_get_num_devices:: Get number of devices for the given device
1846 * acc_set_device_type:: Set type of device accelerator to use.
1847 * acc_get_device_type:: Get type of device accelerator to be used.
1848 * acc_set_device_num:: Set device number to use.
1849 * acc_get_device_num:: Get device number to be used.
1850 * acc_get_property:: Get device property.
1851 * acc_async_test:: Tests for completion of a specific asynchronous
1853 * acc_async_test_all:: Tests for completion of all asynchronous
1855 * acc_wait:: Wait for completion of a specific asynchronous
1857 * acc_wait_all:: Waits for completion of all asynchronous
1859 * acc_wait_all_async:: Wait for completion of all asynchronous
1861 * acc_wait_async:: Wait for completion of asynchronous operations.
1862 * acc_init:: Initialize runtime for a specific device type.
1863 * acc_shutdown:: Shuts down the runtime for a specific device
1865 * acc_on_device:: Whether executing on a particular device
1866 * acc_malloc:: Allocate device memory.
1867 * acc_free:: Free device memory.
1868 * acc_copyin:: Allocate device memory and copy host memory to
1870 * acc_present_or_copyin:: If the data is not present on the device,
1871 allocate device memory and copy from host
1873 * acc_create:: Allocate device memory and map it to host
1875 * acc_present_or_create:: If the data is not present on the device,
1876 allocate device memory and map it to host
1878 * acc_copyout:: Copy device memory to host memory.
1879 * acc_delete:: Free device memory.
1880 * acc_update_device:: Update device memory from mapped host memory.
1881 * acc_update_self:: Update host memory from mapped device memory.
1882 * acc_map_data:: Map previously allocated device memory to host
1884 * acc_unmap_data:: Unmap device memory from host memory.
1885 * acc_deviceptr:: Get device pointer associated with specific
1887 * acc_hostptr:: Get host pointer associated with specific
1889 * acc_is_present:: Indicate whether host variable / array is
1891 * acc_memcpy_to_device:: Copy host memory to device memory.
1892 * acc_memcpy_from_device:: Copy device memory to host memory.
1893 * acc_attach:: Let device pointer point to device-pointer target.
1894 * acc_detach:: Let device pointer point to host-pointer target.
1896 API routines for target platforms.
1898 * acc_get_current_cuda_device:: Get CUDA device handle.
1899 * acc_get_current_cuda_context::Get CUDA context handle.
1900 * acc_get_cuda_stream:: Get CUDA stream handle.
1901 * acc_set_cuda_stream:: Set CUDA stream handle.
1903 API routines for the OpenACC Profiling Interface.
1905 * acc_prof_register:: Register callbacks.
1906 * acc_prof_unregister:: Unregister callbacks.
1907 * acc_prof_lookup:: Obtain inquiry functions.
1908 * acc_register_library:: Library registration.
1913 @node acc_get_num_devices
1914 @section @code{acc_get_num_devices} -- Get number of devices for given device type
1916 @item @emph{Description}
1917 This function returns a value indicating the number of devices available
1918 for the device type specified in @var{devicetype}.
1921 @multitable @columnfractions .20 .80
1922 @item @emph{Prototype}: @tab @code{int acc_get_num_devices(acc_device_t devicetype);}
1925 @item @emph{Fortran}:
1926 @multitable @columnfractions .20 .80
1927 @item @emph{Interface}: @tab @code{integer function acc_get_num_devices(devicetype)}
1928 @item @tab @code{integer(kind=acc_device_kind) devicetype}
1931 @item @emph{Reference}:
1932 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
1938 @node acc_set_device_type
1939 @section @code{acc_set_device_type} -- Set type of device accelerator to use.
1941 @item @emph{Description}
1942 This function indicates to the runtime library which device type, specified
1943 in @var{devicetype}, to use when executing a parallel or kernels region.
1946 @multitable @columnfractions .20 .80
1947 @item @emph{Prototype}: @tab @code{acc_set_device_type(acc_device_t devicetype);}
1950 @item @emph{Fortran}:
1951 @multitable @columnfractions .20 .80
1952 @item @emph{Interface}: @tab @code{subroutine acc_set_device_type(devicetype)}
1953 @item @tab @code{integer(kind=acc_device_kind) devicetype}
1956 @item @emph{Reference}:
1957 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
1963 @node acc_get_device_type
1964 @section @code{acc_get_device_type} -- Get type of device accelerator to be used.
1966 @item @emph{Description}
1967 This function returns what device type will be used when executing a
1968 parallel or kernels region.
1970 This function returns @code{acc_device_none} if
1971 @code{acc_get_device_type} is called from
1972 @code{acc_ev_device_init_start}, @code{acc_ev_device_init_end}
1973 callbacks of the OpenACC Profiling Interface (@ref{OpenACC Profiling
1974 Interface}), that is, if the device is currently being initialized.
1977 @multitable @columnfractions .20 .80
1978 @item @emph{Prototype}: @tab @code{acc_device_t acc_get_device_type(void);}
1981 @item @emph{Fortran}:
1982 @multitable @columnfractions .20 .80
1983 @item @emph{Interface}: @tab @code{function acc_get_device_type(void)}
1984 @item @tab @code{integer(kind=acc_device_kind) acc_get_device_type}
1987 @item @emph{Reference}:
1988 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
1994 @node acc_set_device_num
1995 @section @code{acc_set_device_num} -- Set device number to use.
1997 @item @emph{Description}
1998 This function will indicate to the runtime which device number,
1999 specified by @var{devicenum}, associated with the specified device
2000 type @var{devicetype}.
2003 @multitable @columnfractions .20 .80
2004 @item @emph{Prototype}: @tab @code{acc_set_device_num(int devicenum, acc_device_t devicetype);}
2007 @item @emph{Fortran}:
2008 @multitable @columnfractions .20 .80
2009 @item @emph{Interface}: @tab @code{subroutine acc_set_device_num(devicenum, devicetype)}
2010 @item @tab @code{integer devicenum}
2011 @item @tab @code{integer(kind=acc_device_kind) devicetype}
2014 @item @emph{Reference}:
2015 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2021 @node acc_get_device_num
2022 @section @code{acc_get_device_num} -- Get device number to be used.
2024 @item @emph{Description}
2025 This function returns which device number associated with the specified device
2026 type @var{devicetype}, will be used when executing a parallel or kernels
2030 @multitable @columnfractions .20 .80
2031 @item @emph{Prototype}: @tab @code{int acc_get_device_num(acc_device_t devicetype);}
2034 @item @emph{Fortran}:
2035 @multitable @columnfractions .20 .80
2036 @item @emph{Interface}: @tab @code{function acc_get_device_num(devicetype)}
2037 @item @tab @code{integer(kind=acc_device_kind) devicetype}
2038 @item @tab @code{integer acc_get_device_num}
2041 @item @emph{Reference}:
2042 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2048 @node acc_get_property
2049 @section @code{acc_get_property} -- Get device property.
2050 @cindex acc_get_property
2051 @cindex acc_get_property_string
2053 @item @emph{Description}
2054 These routines return the value of the specified @var{property} for the
2055 device being queried according to @var{devicenum} and @var{devicetype}.
2056 Integer-valued and string-valued properties are returned by
2057 @code{acc_get_property} and @code{acc_get_property_string} respectively.
2058 The Fortran @code{acc_get_property_string} subroutine returns the string
2059 retrieved in its fourth argument while the remaining entry points are
2060 functions, which pass the return value as their result.
2062 Note for Fortran, only: the OpenACC technical committee corrected and, hence,
2063 modified the interface introduced in OpenACC 2.6. The kind-value parameter
2064 @code{acc_device_property} has been renamed to @code{acc_device_property_kind}
2065 for consistency and the return type of the @code{acc_get_property} function is
2066 now a @code{c_size_t} integer instead of a @code{acc_device_property} integer.
2067 The parameter @code{acc_device_property} will continue to be provided,
2068 but might be removed in a future version of GCC.
2071 @multitable @columnfractions .20 .80
2072 @item @emph{Prototype}: @tab @code{size_t acc_get_property(int devicenum, acc_device_t devicetype, acc_device_property_t property);}
2073 @item @emph{Prototype}: @tab @code{const char *acc_get_property_string(int devicenum, acc_device_t devicetype, acc_device_property_t property);}
2076 @item @emph{Fortran}:
2077 @multitable @columnfractions .20 .80
2078 @item @emph{Interface}: @tab @code{function acc_get_property(devicenum, devicetype, property)}
2079 @item @emph{Interface}: @tab @code{subroutine acc_get_property_string(devicenum, devicetype, property, string)}
2080 @item @tab @code{use ISO_C_Binding, only: c_size_t}
2081 @item @tab @code{integer devicenum}
2082 @item @tab @code{integer(kind=acc_device_kind) devicetype}
2083 @item @tab @code{integer(kind=acc_device_property_kind) property}
2084 @item @tab @code{integer(kind=c_size_t) acc_get_property}
2085 @item @tab @code{character(*) string}
2088 @item @emph{Reference}:
2089 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2095 @node acc_async_test
2096 @section @code{acc_async_test} -- Test for completion of a specific asynchronous operation.
2098 @item @emph{Description}
2099 This function tests for completion of the asynchronous operation specified
2100 in @var{arg}. In C/C++, a non-zero value will be returned to indicate
2101 the specified asynchronous operation has completed. While Fortran will return
2102 a @code{true}. If the asynchronous operation has not completed, C/C++ returns
2103 a zero and Fortran returns a @code{false}.
2106 @multitable @columnfractions .20 .80
2107 @item @emph{Prototype}: @tab @code{int acc_async_test(int arg);}
2110 @item @emph{Fortran}:
2111 @multitable @columnfractions .20 .80
2112 @item @emph{Interface}: @tab @code{function acc_async_test(arg)}
2113 @item @tab @code{integer(kind=acc_handle_kind) arg}
2114 @item @tab @code{logical acc_async_test}
2117 @item @emph{Reference}:
2118 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2124 @node acc_async_test_all
2125 @section @code{acc_async_test_all} -- Tests for completion of all asynchronous operations.
2127 @item @emph{Description}
2128 This function tests for completion of all asynchronous operations.
2129 In C/C++, a non-zero value will be returned to indicate all asynchronous
2130 operations have completed. While Fortran will return a @code{true}. If
2131 any asynchronous operation has not completed, C/C++ returns a zero and
2132 Fortran returns a @code{false}.
2135 @multitable @columnfractions .20 .80
2136 @item @emph{Prototype}: @tab @code{int acc_async_test_all(void);}
2139 @item @emph{Fortran}:
2140 @multitable @columnfractions .20 .80
2141 @item @emph{Interface}: @tab @code{function acc_async_test()}
2142 @item @tab @code{logical acc_get_device_num}
2145 @item @emph{Reference}:
2146 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2153 @section @code{acc_wait} -- Wait for completion of a specific asynchronous operation.
2155 @item @emph{Description}
2156 This function waits for completion of the asynchronous operation
2157 specified in @var{arg}.
2160 @multitable @columnfractions .20 .80
2161 @item @emph{Prototype}: @tab @code{acc_wait(arg);}
2162 @item @emph{Prototype (OpenACC 1.0 compatibility)}: @tab @code{acc_async_wait(arg);}
2165 @item @emph{Fortran}:
2166 @multitable @columnfractions .20 .80
2167 @item @emph{Interface}: @tab @code{subroutine acc_wait(arg)}
2168 @item @tab @code{integer(acc_handle_kind) arg}
2169 @item @emph{Interface (OpenACC 1.0 compatibility)}: @tab @code{subroutine acc_async_wait(arg)}
2170 @item @tab @code{integer(acc_handle_kind) arg}
2173 @item @emph{Reference}:
2174 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2181 @section @code{acc_wait_all} -- Waits for completion of all asynchronous operations.
2183 @item @emph{Description}
2184 This function waits for the completion of all asynchronous operations.
2187 @multitable @columnfractions .20 .80
2188 @item @emph{Prototype}: @tab @code{acc_wait_all(void);}
2189 @item @emph{Prototype (OpenACC 1.0 compatibility)}: @tab @code{acc_async_wait_all(void);}
2192 @item @emph{Fortran}:
2193 @multitable @columnfractions .20 .80
2194 @item @emph{Interface}: @tab @code{subroutine acc_wait_all()}
2195 @item @emph{Interface (OpenACC 1.0 compatibility)}: @tab @code{subroutine acc_async_wait_all()}
2198 @item @emph{Reference}:
2199 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2205 @node acc_wait_all_async
2206 @section @code{acc_wait_all_async} -- Wait for completion of all asynchronous operations.
2208 @item @emph{Description}
2209 This function enqueues a wait operation on the queue @var{async} for any
2210 and all asynchronous operations that have been previously enqueued on
2214 @multitable @columnfractions .20 .80
2215 @item @emph{Prototype}: @tab @code{acc_wait_all_async(int async);}
2218 @item @emph{Fortran}:
2219 @multitable @columnfractions .20 .80
2220 @item @emph{Interface}: @tab @code{subroutine acc_wait_all_async(async)}
2221 @item @tab @code{integer(acc_handle_kind) async}
2224 @item @emph{Reference}:
2225 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2231 @node acc_wait_async
2232 @section @code{acc_wait_async} -- Wait for completion of asynchronous operations.
2234 @item @emph{Description}
2235 This function enqueues a wait operation on queue @var{async} for any and all
2236 asynchronous operations enqueued on queue @var{arg}.
2239 @multitable @columnfractions .20 .80
2240 @item @emph{Prototype}: @tab @code{acc_wait_async(int arg, int async);}
2243 @item @emph{Fortran}:
2244 @multitable @columnfractions .20 .80
2245 @item @emph{Interface}: @tab @code{subroutine acc_wait_async(arg, async)}
2246 @item @tab @code{integer(acc_handle_kind) arg, async}
2249 @item @emph{Reference}:
2250 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2257 @section @code{acc_init} -- Initialize runtime for a specific device type.
2259 @item @emph{Description}
2260 This function initializes the runtime for the device type specified in
2264 @multitable @columnfractions .20 .80
2265 @item @emph{Prototype}: @tab @code{acc_init(acc_device_t devicetype);}
2268 @item @emph{Fortran}:
2269 @multitable @columnfractions .20 .80
2270 @item @emph{Interface}: @tab @code{subroutine acc_init(devicetype)}
2271 @item @tab @code{integer(acc_device_kind) devicetype}
2274 @item @emph{Reference}:
2275 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2282 @section @code{acc_shutdown} -- Shuts down the runtime for a specific device type.
2284 @item @emph{Description}
2285 This function shuts down the runtime for the device type specified in
2289 @multitable @columnfractions .20 .80
2290 @item @emph{Prototype}: @tab @code{acc_shutdown(acc_device_t devicetype);}
2293 @item @emph{Fortran}:
2294 @multitable @columnfractions .20 .80
2295 @item @emph{Interface}: @tab @code{subroutine acc_shutdown(devicetype)}
2296 @item @tab @code{integer(acc_device_kind) devicetype}
2299 @item @emph{Reference}:
2300 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2307 @section @code{acc_on_device} -- Whether executing on a particular device
2309 @item @emph{Description}:
2310 This function returns whether the program is executing on a particular
2311 device specified in @var{devicetype}. In C/C++ a non-zero value is
2312 returned to indicate the device is executing on the specified device type.
2313 In Fortran, @code{true} will be returned. If the program is not executing
2314 on the specified device type C/C++ will return a zero, while Fortran will
2315 return @code{false}.
2318 @multitable @columnfractions .20 .80
2319 @item @emph{Prototype}: @tab @code{acc_on_device(acc_device_t devicetype);}
2322 @item @emph{Fortran}:
2323 @multitable @columnfractions .20 .80
2324 @item @emph{Interface}: @tab @code{function acc_on_device(devicetype)}
2325 @item @tab @code{integer(acc_device_kind) devicetype}
2326 @item @tab @code{logical acc_on_device}
2330 @item @emph{Reference}:
2331 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2338 @section @code{acc_malloc} -- Allocate device memory.
2340 @item @emph{Description}
2341 This function allocates @var{len} bytes of device memory. It returns
2342 the device address of the allocated memory.
2345 @multitable @columnfractions .20 .80
2346 @item @emph{Prototype}: @tab @code{d_void* acc_malloc(size_t len);}
2349 @item @emph{Reference}:
2350 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2357 @section @code{acc_free} -- Free device memory.
2359 @item @emph{Description}
2360 Free previously allocated device memory at the device address @code{a}.
2363 @multitable @columnfractions .20 .80
2364 @item @emph{Prototype}: @tab @code{acc_free(d_void *a);}
2367 @item @emph{Reference}:
2368 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2375 @section @code{acc_copyin} -- Allocate device memory and copy host memory to it.
2377 @item @emph{Description}
2378 In C/C++, this function allocates @var{len} bytes of device memory
2379 and maps it to the specified host address in @var{a}. The device
2380 address of the newly allocated device memory is returned.
2382 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
2383 a contiguous array section. The second form @var{a} specifies a
2384 variable or array element and @var{len} specifies the length in bytes.
2387 @multitable @columnfractions .20 .80
2388 @item @emph{Prototype}: @tab @code{void *acc_copyin(h_void *a, size_t len);}
2389 @item @emph{Prototype}: @tab @code{void *acc_copyin_async(h_void *a, size_t len, int async);}
2392 @item @emph{Fortran}:
2393 @multitable @columnfractions .20 .80
2394 @item @emph{Interface}: @tab @code{subroutine acc_copyin(a)}
2395 @item @tab @code{type, dimension(:[,:]...) :: a}
2396 @item @emph{Interface}: @tab @code{subroutine acc_copyin(a, len)}
2397 @item @tab @code{type, dimension(:[,:]...) :: a}
2398 @item @tab @code{integer len}
2399 @item @emph{Interface}: @tab @code{subroutine acc_copyin_async(a, async)}
2400 @item @tab @code{type, dimension(:[,:]...) :: a}
2401 @item @tab @code{integer(acc_handle_kind) :: async}
2402 @item @emph{Interface}: @tab @code{subroutine acc_copyin_async(a, len, async)}
2403 @item @tab @code{type, dimension(:[,:]...) :: a}
2404 @item @tab @code{integer len}
2405 @item @tab @code{integer(acc_handle_kind) :: async}
2408 @item @emph{Reference}:
2409 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2415 @node acc_present_or_copyin
2416 @section @code{acc_present_or_copyin} -- If the data is not present on the device, allocate device memory and copy from host memory.
2418 @item @emph{Description}
2419 This function tests if the host data specified by @var{a} and of length
2420 @var{len} is present or not. If it is not present, then device memory
2421 will be allocated and the host memory copied. The device address of
2422 the newly allocated device memory is returned.
2424 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
2425 a contiguous array section. The second form @var{a} specifies a variable or
2426 array element and @var{len} specifies the length in bytes.
2428 Note that @code{acc_present_or_copyin} and @code{acc_pcopyin} exist for
2429 backward compatibility with OpenACC 2.0; use @ref{acc_copyin} instead.
2432 @multitable @columnfractions .20 .80
2433 @item @emph{Prototype}: @tab @code{void *acc_present_or_copyin(h_void *a, size_t len);}
2434 @item @emph{Prototype}: @tab @code{void *acc_pcopyin(h_void *a, size_t len);}
2437 @item @emph{Fortran}:
2438 @multitable @columnfractions .20 .80
2439 @item @emph{Interface}: @tab @code{subroutine acc_present_or_copyin(a)}
2440 @item @tab @code{type, dimension(:[,:]...) :: a}
2441 @item @emph{Interface}: @tab @code{subroutine acc_present_or_copyin(a, len)}
2442 @item @tab @code{type, dimension(:[,:]...) :: a}
2443 @item @tab @code{integer len}
2444 @item @emph{Interface}: @tab @code{subroutine acc_pcopyin(a)}
2445 @item @tab @code{type, dimension(:[,:]...) :: a}
2446 @item @emph{Interface}: @tab @code{subroutine acc_pcopyin(a, len)}
2447 @item @tab @code{type, dimension(:[,:]...) :: a}
2448 @item @tab @code{integer len}
2451 @item @emph{Reference}:
2452 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2459 @section @code{acc_create} -- Allocate device memory and map it to host memory.
2461 @item @emph{Description}
2462 This function allocates device memory and maps it to host memory specified
2463 by the host address @var{a} with a length of @var{len} bytes. In C/C++,
2464 the function returns the device address of the allocated device memory.
2466 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
2467 a contiguous array section. The second form @var{a} specifies a variable or
2468 array element and @var{len} specifies the length in bytes.
2471 @multitable @columnfractions .20 .80
2472 @item @emph{Prototype}: @tab @code{void *acc_create(h_void *a, size_t len);}
2473 @item @emph{Prototype}: @tab @code{void *acc_create_async(h_void *a, size_t len, int async);}
2476 @item @emph{Fortran}:
2477 @multitable @columnfractions .20 .80
2478 @item @emph{Interface}: @tab @code{subroutine acc_create(a)}
2479 @item @tab @code{type, dimension(:[,:]...) :: a}
2480 @item @emph{Interface}: @tab @code{subroutine acc_create(a, len)}
2481 @item @tab @code{type, dimension(:[,:]...) :: a}
2482 @item @tab @code{integer len}
2483 @item @emph{Interface}: @tab @code{subroutine acc_create_async(a, async)}
2484 @item @tab @code{type, dimension(:[,:]...) :: a}
2485 @item @tab @code{integer(acc_handle_kind) :: async}
2486 @item @emph{Interface}: @tab @code{subroutine acc_create_async(a, len, async)}
2487 @item @tab @code{type, dimension(:[,:]...) :: a}
2488 @item @tab @code{integer len}
2489 @item @tab @code{integer(acc_handle_kind) :: async}
2492 @item @emph{Reference}:
2493 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2499 @node acc_present_or_create
2500 @section @code{acc_present_or_create} -- If the data is not present on the device, allocate device memory and map it to host memory.
2502 @item @emph{Description}
2503 This function tests if the host data specified by @var{a} and of length
2504 @var{len} is present or not. If it is not present, then device memory
2505 will be allocated and mapped to host memory. In C/C++, the device address
2506 of the newly allocated device memory is returned.
2508 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
2509 a contiguous array section. The second form @var{a} specifies a variable or
2510 array element and @var{len} specifies the length in bytes.
2512 Note that @code{acc_present_or_create} and @code{acc_pcreate} exist for
2513 backward compatibility with OpenACC 2.0; use @ref{acc_create} instead.
2516 @multitable @columnfractions .20 .80
2517 @item @emph{Prototype}: @tab @code{void *acc_present_or_create(h_void *a, size_t len)}
2518 @item @emph{Prototype}: @tab @code{void *acc_pcreate(h_void *a, size_t len)}
2521 @item @emph{Fortran}:
2522 @multitable @columnfractions .20 .80
2523 @item @emph{Interface}: @tab @code{subroutine acc_present_or_create(a)}
2524 @item @tab @code{type, dimension(:[,:]...) :: a}
2525 @item @emph{Interface}: @tab @code{subroutine acc_present_or_create(a, len)}
2526 @item @tab @code{type, dimension(:[,:]...) :: a}
2527 @item @tab @code{integer len}
2528 @item @emph{Interface}: @tab @code{subroutine acc_pcreate(a)}
2529 @item @tab @code{type, dimension(:[,:]...) :: a}
2530 @item @emph{Interface}: @tab @code{subroutine acc_pcreate(a, len)}
2531 @item @tab @code{type, dimension(:[,:]...) :: a}
2532 @item @tab @code{integer len}
2535 @item @emph{Reference}:
2536 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2543 @section @code{acc_copyout} -- Copy device memory to host memory.
2545 @item @emph{Description}
2546 This function copies mapped device memory to host memory which is specified
2547 by host address @var{a} for a length @var{len} bytes in C/C++.
2549 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
2550 a contiguous array section. The second form @var{a} specifies a variable or
2551 array element and @var{len} specifies the length in bytes.
2554 @multitable @columnfractions .20 .80
2555 @item @emph{Prototype}: @tab @code{acc_copyout(h_void *a, size_t len);}
2556 @item @emph{Prototype}: @tab @code{acc_copyout_async(h_void *a, size_t len, int async);}
2557 @item @emph{Prototype}: @tab @code{acc_copyout_finalize(h_void *a, size_t len);}
2558 @item @emph{Prototype}: @tab @code{acc_copyout_finalize_async(h_void *a, size_t len, int async);}
2561 @item @emph{Fortran}:
2562 @multitable @columnfractions .20 .80
2563 @item @emph{Interface}: @tab @code{subroutine acc_copyout(a)}
2564 @item @tab @code{type, dimension(:[,:]...) :: a}
2565 @item @emph{Interface}: @tab @code{subroutine acc_copyout(a, len)}
2566 @item @tab @code{type, dimension(:[,:]...) :: a}
2567 @item @tab @code{integer len}
2568 @item @emph{Interface}: @tab @code{subroutine acc_copyout_async(a, async)}
2569 @item @tab @code{type, dimension(:[,:]...) :: a}
2570 @item @tab @code{integer(acc_handle_kind) :: async}
2571 @item @emph{Interface}: @tab @code{subroutine acc_copyout_async(a, len, async)}
2572 @item @tab @code{type, dimension(:[,:]...) :: a}
2573 @item @tab @code{integer len}
2574 @item @tab @code{integer(acc_handle_kind) :: async}
2575 @item @emph{Interface}: @tab @code{subroutine acc_copyout_finalize(a)}
2576 @item @tab @code{type, dimension(:[,:]...) :: a}
2577 @item @emph{Interface}: @tab @code{subroutine acc_copyout_finalize(a, len)}
2578 @item @tab @code{type, dimension(:[,:]...) :: a}
2579 @item @tab @code{integer len}
2580 @item @emph{Interface}: @tab @code{subroutine acc_copyout_finalize_async(a, async)}
2581 @item @tab @code{type, dimension(:[,:]...) :: a}
2582 @item @tab @code{integer(acc_handle_kind) :: async}
2583 @item @emph{Interface}: @tab @code{subroutine acc_copyout_finalize_async(a, len, async)}
2584 @item @tab @code{type, dimension(:[,:]...) :: a}
2585 @item @tab @code{integer len}
2586 @item @tab @code{integer(acc_handle_kind) :: async}
2589 @item @emph{Reference}:
2590 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2597 @section @code{acc_delete} -- Free device memory.
2599 @item @emph{Description}
2600 This function frees previously allocated device memory specified by
2601 the device address @var{a} and the length of @var{len} bytes.
2603 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
2604 a contiguous array section. The second form @var{a} specifies a variable or
2605 array element and @var{len} specifies the length in bytes.
2608 @multitable @columnfractions .20 .80
2609 @item @emph{Prototype}: @tab @code{acc_delete(h_void *a, size_t len);}
2610 @item @emph{Prototype}: @tab @code{acc_delete_async(h_void *a, size_t len, int async);}
2611 @item @emph{Prototype}: @tab @code{acc_delete_finalize(h_void *a, size_t len);}
2612 @item @emph{Prototype}: @tab @code{acc_delete_finalize_async(h_void *a, size_t len, int async);}
2615 @item @emph{Fortran}:
2616 @multitable @columnfractions .20 .80
2617 @item @emph{Interface}: @tab @code{subroutine acc_delete(a)}
2618 @item @tab @code{type, dimension(:[,:]...) :: a}
2619 @item @emph{Interface}: @tab @code{subroutine acc_delete(a, len)}
2620 @item @tab @code{type, dimension(:[,:]...) :: a}
2621 @item @tab @code{integer len}
2622 @item @emph{Interface}: @tab @code{subroutine acc_delete_async(a, async)}
2623 @item @tab @code{type, dimension(:[,:]...) :: a}
2624 @item @tab @code{integer(acc_handle_kind) :: async}
2625 @item @emph{Interface}: @tab @code{subroutine acc_delete_async(a, len, async)}
2626 @item @tab @code{type, dimension(:[,:]...) :: a}
2627 @item @tab @code{integer len}
2628 @item @tab @code{integer(acc_handle_kind) :: async}
2629 @item @emph{Interface}: @tab @code{subroutine acc_delete_finalize(a)}
2630 @item @tab @code{type, dimension(:[,:]...) :: a}
2631 @item @emph{Interface}: @tab @code{subroutine acc_delete_finalize(a, len)}
2632 @item @tab @code{type, dimension(:[,:]...) :: a}
2633 @item @tab @code{integer len}
2634 @item @emph{Interface}: @tab @code{subroutine acc_delete_async_finalize(a, async)}
2635 @item @tab @code{type, dimension(:[,:]...) :: a}
2636 @item @tab @code{integer(acc_handle_kind) :: async}
2637 @item @emph{Interface}: @tab @code{subroutine acc_delete_async_finalize(a, len, async)}
2638 @item @tab @code{type, dimension(:[,:]...) :: a}
2639 @item @tab @code{integer len}
2640 @item @tab @code{integer(acc_handle_kind) :: async}
2643 @item @emph{Reference}:
2644 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2650 @node acc_update_device
2651 @section @code{acc_update_device} -- Update device memory from mapped host memory.
2653 @item @emph{Description}
2654 This function updates the device copy from the previously mapped host memory.
2655 The host memory is specified with the host address @var{a} and a length of
2658 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
2659 a contiguous array section. The second form @var{a} specifies a variable or
2660 array element and @var{len} specifies the length in bytes.
2663 @multitable @columnfractions .20 .80
2664 @item @emph{Prototype}: @tab @code{acc_update_device(h_void *a, size_t len);}
2665 @item @emph{Prototype}: @tab @code{acc_update_device(h_void *a, size_t len, async);}
2668 @item @emph{Fortran}:
2669 @multitable @columnfractions .20 .80
2670 @item @emph{Interface}: @tab @code{subroutine acc_update_device(a)}
2671 @item @tab @code{type, dimension(:[,:]...) :: a}
2672 @item @emph{Interface}: @tab @code{subroutine acc_update_device(a, len)}
2673 @item @tab @code{type, dimension(:[,:]...) :: a}
2674 @item @tab @code{integer len}
2675 @item @emph{Interface}: @tab @code{subroutine acc_update_device_async(a, async)}
2676 @item @tab @code{type, dimension(:[,:]...) :: a}
2677 @item @tab @code{integer(acc_handle_kind) :: async}
2678 @item @emph{Interface}: @tab @code{subroutine acc_update_device_async(a, len, async)}
2679 @item @tab @code{type, dimension(:[,:]...) :: a}
2680 @item @tab @code{integer len}
2681 @item @tab @code{integer(acc_handle_kind) :: async}
2684 @item @emph{Reference}:
2685 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2691 @node acc_update_self
2692 @section @code{acc_update_self} -- Update host memory from mapped device memory.
2694 @item @emph{Description}
2695 This function updates the host copy from the previously mapped device memory.
2696 The host memory is specified with the host address @var{a} and a length of
2699 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
2700 a contiguous array section. The second form @var{a} specifies a variable or
2701 array element and @var{len} specifies the length in bytes.
2704 @multitable @columnfractions .20 .80
2705 @item @emph{Prototype}: @tab @code{acc_update_self(h_void *a, size_t len);}
2706 @item @emph{Prototype}: @tab @code{acc_update_self_async(h_void *a, size_t len, int async);}
2709 @item @emph{Fortran}:
2710 @multitable @columnfractions .20 .80
2711 @item @emph{Interface}: @tab @code{subroutine acc_update_self(a)}
2712 @item @tab @code{type, dimension(:[,:]...) :: a}
2713 @item @emph{Interface}: @tab @code{subroutine acc_update_self(a, len)}
2714 @item @tab @code{type, dimension(:[,:]...) :: a}
2715 @item @tab @code{integer len}
2716 @item @emph{Interface}: @tab @code{subroutine acc_update_self_async(a, async)}
2717 @item @tab @code{type, dimension(:[,:]...) :: a}
2718 @item @tab @code{integer(acc_handle_kind) :: async}
2719 @item @emph{Interface}: @tab @code{subroutine acc_update_self_async(a, len, async)}
2720 @item @tab @code{type, dimension(:[,:]...) :: a}
2721 @item @tab @code{integer len}
2722 @item @tab @code{integer(acc_handle_kind) :: async}
2725 @item @emph{Reference}:
2726 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2733 @section @code{acc_map_data} -- Map previously allocated device memory to host memory.
2735 @item @emph{Description}
2736 This function maps previously allocated device and host memory. The device
2737 memory is specified with the device address @var{d}. The host memory is
2738 specified with the host address @var{h} and a length of @var{len}.
2741 @multitable @columnfractions .20 .80
2742 @item @emph{Prototype}: @tab @code{acc_map_data(h_void *h, d_void *d, size_t len);}
2745 @item @emph{Reference}:
2746 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2752 @node acc_unmap_data
2753 @section @code{acc_unmap_data} -- Unmap device memory from host memory.
2755 @item @emph{Description}
2756 This function unmaps previously mapped device and host memory. The latter
2757 specified by @var{h}.
2760 @multitable @columnfractions .20 .80
2761 @item @emph{Prototype}: @tab @code{acc_unmap_data(h_void *h);}
2764 @item @emph{Reference}:
2765 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2772 @section @code{acc_deviceptr} -- Get device pointer associated with specific host address.
2774 @item @emph{Description}
2775 This function returns the device address that has been mapped to the
2776 host address specified by @var{h}.
2779 @multitable @columnfractions .20 .80
2780 @item @emph{Prototype}: @tab @code{void *acc_deviceptr(h_void *h);}
2783 @item @emph{Reference}:
2784 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2791 @section @code{acc_hostptr} -- Get host pointer associated with specific device address.
2793 @item @emph{Description}
2794 This function returns the host address that has been mapped to the
2795 device address specified by @var{d}.
2798 @multitable @columnfractions .20 .80
2799 @item @emph{Prototype}: @tab @code{void *acc_hostptr(d_void *d);}
2802 @item @emph{Reference}:
2803 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2809 @node acc_is_present
2810 @section @code{acc_is_present} -- Indicate whether host variable / array is present on device.
2812 @item @emph{Description}
2813 This function indicates whether the specified host address in @var{a} and a
2814 length of @var{len} bytes is present on the device. In C/C++, a non-zero
2815 value is returned to indicate the presence of the mapped memory on the
2816 device. A zero is returned to indicate the memory is not mapped on the
2819 In Fortran, two (2) forms are supported. In the first form, @var{a} specifies
2820 a contiguous array section. The second form @var{a} specifies a variable or
2821 array element and @var{len} specifies the length in bytes. If the host
2822 memory is mapped to device memory, then a @code{true} is returned. Otherwise,
2823 a @code{false} is return to indicate the mapped memory is not present.
2826 @multitable @columnfractions .20 .80
2827 @item @emph{Prototype}: @tab @code{int acc_is_present(h_void *a, size_t len);}
2830 @item @emph{Fortran}:
2831 @multitable @columnfractions .20 .80
2832 @item @emph{Interface}: @tab @code{function acc_is_present(a)}
2833 @item @tab @code{type, dimension(:[,:]...) :: a}
2834 @item @tab @code{logical acc_is_present}
2835 @item @emph{Interface}: @tab @code{function acc_is_present(a, len)}
2836 @item @tab @code{type, dimension(:[,:]...) :: a}
2837 @item @tab @code{integer len}
2838 @item @tab @code{logical acc_is_present}
2841 @item @emph{Reference}:
2842 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2848 @node acc_memcpy_to_device
2849 @section @code{acc_memcpy_to_device} -- Copy host memory to device memory.
2851 @item @emph{Description}
2852 This function copies host memory specified by host address of @var{src} to
2853 device memory specified by the device address @var{dest} for a length of
2857 @multitable @columnfractions .20 .80
2858 @item @emph{Prototype}: @tab @code{acc_memcpy_to_device(d_void *dest, h_void *src, size_t bytes);}
2861 @item @emph{Reference}:
2862 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2868 @node acc_memcpy_from_device
2869 @section @code{acc_memcpy_from_device} -- Copy device memory to host memory.
2871 @item @emph{Description}
2872 This function copies host memory specified by host address of @var{src} from
2873 device memory specified by the device address @var{dest} for a length of
2877 @multitable @columnfractions .20 .80
2878 @item @emph{Prototype}: @tab @code{acc_memcpy_from_device(d_void *dest, h_void *src, size_t bytes);}
2881 @item @emph{Reference}:
2882 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2889 @section @code{acc_attach} -- Let device pointer point to device-pointer target.
2891 @item @emph{Description}
2892 This function updates a pointer on the device from pointing to a host-pointer
2893 address to pointing to the corresponding device data.
2896 @multitable @columnfractions .20 .80
2897 @item @emph{Prototype}: @tab @code{acc_attach(h_void **ptr);}
2898 @item @emph{Prototype}: @tab @code{acc_attach_async(h_void **ptr, int async);}
2901 @item @emph{Reference}:
2902 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2909 @section @code{acc_detach} -- Let device pointer point to host-pointer target.
2911 @item @emph{Description}
2912 This function updates a pointer on the device from pointing to a device-pointer
2913 address to pointing to the corresponding host data.
2916 @multitable @columnfractions .20 .80
2917 @item @emph{Prototype}: @tab @code{acc_detach(h_void **ptr);}
2918 @item @emph{Prototype}: @tab @code{acc_detach_async(h_void **ptr, int async);}
2919 @item @emph{Prototype}: @tab @code{acc_detach_finalize(h_void **ptr);}
2920 @item @emph{Prototype}: @tab @code{acc_detach_finalize_async(h_void **ptr, int async);}
2923 @item @emph{Reference}:
2924 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2930 @node acc_get_current_cuda_device
2931 @section @code{acc_get_current_cuda_device} -- Get CUDA device handle.
2933 @item @emph{Description}
2934 This function returns the CUDA device handle. This handle is the same
2935 as used by the CUDA Runtime or Driver API's.
2938 @multitable @columnfractions .20 .80
2939 @item @emph{Prototype}: @tab @code{void *acc_get_current_cuda_device(void);}
2942 @item @emph{Reference}:
2943 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2949 @node acc_get_current_cuda_context
2950 @section @code{acc_get_current_cuda_context} -- Get CUDA context handle.
2952 @item @emph{Description}
2953 This function returns the CUDA context handle. This handle is the same
2954 as used by the CUDA Runtime or Driver API's.
2957 @multitable @columnfractions .20 .80
2958 @item @emph{Prototype}: @tab @code{void *acc_get_current_cuda_context(void);}
2961 @item @emph{Reference}:
2962 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2968 @node acc_get_cuda_stream
2969 @section @code{acc_get_cuda_stream} -- Get CUDA stream handle.
2971 @item @emph{Description}
2972 This function returns the CUDA stream handle for the queue @var{async}.
2973 This handle is the same as used by the CUDA Runtime or Driver API's.
2976 @multitable @columnfractions .20 .80
2977 @item @emph{Prototype}: @tab @code{void *acc_get_cuda_stream(int async);}
2980 @item @emph{Reference}:
2981 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
2987 @node acc_set_cuda_stream
2988 @section @code{acc_set_cuda_stream} -- Set CUDA stream handle.
2990 @item @emph{Description}
2991 This function associates the stream handle specified by @var{stream} with
2992 the queue @var{async}.
2994 This cannot be used to change the stream handle associated with
2995 @code{acc_async_sync}.
2997 The return value is not specified.
3000 @multitable @columnfractions .20 .80
3001 @item @emph{Prototype}: @tab @code{int acc_set_cuda_stream(int async, void *stream);}
3004 @item @emph{Reference}:
3005 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
3011 @node acc_prof_register
3012 @section @code{acc_prof_register} -- Register callbacks.
3014 @item @emph{Description}:
3015 This function registers callbacks.
3018 @multitable @columnfractions .20 .80
3019 @item @emph{Prototype}: @tab @code{void acc_prof_register (acc_event_t, acc_prof_callback, acc_register_t);}
3022 @item @emph{See also}:
3023 @ref{OpenACC Profiling Interface}
3025 @item @emph{Reference}:
3026 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
3032 @node acc_prof_unregister
3033 @section @code{acc_prof_unregister} -- Unregister callbacks.
3035 @item @emph{Description}:
3036 This function unregisters callbacks.
3039 @multitable @columnfractions .20 .80
3040 @item @emph{Prototype}: @tab @code{void acc_prof_unregister (acc_event_t, acc_prof_callback, acc_register_t);}
3043 @item @emph{See also}:
3044 @ref{OpenACC Profiling Interface}
3046 @item @emph{Reference}:
3047 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
3053 @node acc_prof_lookup
3054 @section @code{acc_prof_lookup} -- Obtain inquiry functions.
3056 @item @emph{Description}:
3057 Function to obtain inquiry functions.
3060 @multitable @columnfractions .20 .80
3061 @item @emph{Prototype}: @tab @code{acc_query_fn acc_prof_lookup (const char *);}
3064 @item @emph{See also}:
3065 @ref{OpenACC Profiling Interface}
3067 @item @emph{Reference}:
3068 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
3074 @node acc_register_library
3075 @section @code{acc_register_library} -- Library registration.
3077 @item @emph{Description}:
3078 Function for library registration.
3081 @multitable @columnfractions .20 .80
3082 @item @emph{Prototype}: @tab @code{void acc_register_library (acc_prof_reg, acc_prof_reg, acc_prof_lookup_func);}
3085 @item @emph{See also}:
3086 @ref{OpenACC Profiling Interface}, @ref{ACC_PROFLIB}
3088 @item @emph{Reference}:
3089 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
3095 @c ---------------------------------------------------------------------
3096 @c OpenACC Environment Variables
3097 @c ---------------------------------------------------------------------
3099 @node OpenACC Environment Variables
3100 @chapter OpenACC Environment Variables
3102 The variables @env{ACC_DEVICE_TYPE} and @env{ACC_DEVICE_NUM}
3103 are defined by section 4 of the OpenACC specification in version 2.0.
3104 The variable @env{ACC_PROFLIB}
3105 is defined by section 4 of the OpenACC specification in version 2.6.
3106 The variable @env{GCC_ACC_NOTIFY} is used for diagnostic purposes.
3117 @node ACC_DEVICE_TYPE
3118 @section @code{ACC_DEVICE_TYPE}
3120 @item @emph{Reference}:
3121 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
3127 @node ACC_DEVICE_NUM
3128 @section @code{ACC_DEVICE_NUM}
3130 @item @emph{Reference}:
3131 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
3138 @section @code{ACC_PROFLIB}
3140 @item @emph{See also}:
3141 @ref{acc_register_library}, @ref{OpenACC Profiling Interface}
3143 @item @emph{Reference}:
3144 @uref{https://www.openacc.org, OpenACC specification v2.6}, section
3150 @node GCC_ACC_NOTIFY
3151 @section @code{GCC_ACC_NOTIFY}
3153 @item @emph{Description}:
3154 Print debug information pertaining to the accelerator.
3159 @c ---------------------------------------------------------------------
3160 @c CUDA Streams Usage
3161 @c ---------------------------------------------------------------------
3163 @node CUDA Streams Usage
3164 @chapter CUDA Streams Usage
3166 This applies to the @code{nvptx} plugin only.
3168 The library provides elements that perform asynchronous movement of
3169 data and asynchronous operation of computing constructs. This
3170 asynchronous functionality is implemented by making use of CUDA
3171 streams@footnote{See "Stream Management" in "CUDA Driver API",
3172 TRM-06703-001, Version 5.5, for additional information}.
3174 The primary means by that the asynchronous functionality is accessed
3175 is through the use of those OpenACC directives which make use of the
3176 @code{async} and @code{wait} clauses. When the @code{async} clause is
3177 first used with a directive, it creates a CUDA stream. If an
3178 @code{async-argument} is used with the @code{async} clause, then the
3179 stream is associated with the specified @code{async-argument}.
3181 Following the creation of an association between a CUDA stream and the
3182 @code{async-argument} of an @code{async} clause, both the @code{wait}
3183 clause and the @code{wait} directive can be used. When either the
3184 clause or directive is used after stream creation, it creates a
3185 rendezvous point whereby execution waits until all operations
3186 associated with the @code{async-argument}, that is, stream, have
3189 Normally, the management of the streams that are created as a result of
3190 using the @code{async} clause, is done without any intervention by the
3191 caller. This implies the association between the @code{async-argument}
3192 and the CUDA stream will be maintained for the lifetime of the program.
3193 However, this association can be changed through the use of the library
3194 function @code{acc_set_cuda_stream}. When the function
3195 @code{acc_set_cuda_stream} is called, the CUDA stream that was
3196 originally associated with the @code{async} clause will be destroyed.
3197 Caution should be taken when changing the association as subsequent
3198 references to the @code{async-argument} refer to a different
3203 @c ---------------------------------------------------------------------
3204 @c OpenACC Library Interoperability
3205 @c ---------------------------------------------------------------------
3207 @node OpenACC Library Interoperability
3208 @chapter OpenACC Library Interoperability
3210 @section Introduction
3212 The OpenACC library uses the CUDA Driver API, and may interact with
3213 programs that use the Runtime library directly, or another library
3214 based on the Runtime library, e.g., CUBLAS@footnote{See section 2.26,
3215 "Interactions with the CUDA Driver API" in
3216 "CUDA Runtime API", Version 5.5, and section 2.27, "VDPAU
3217 Interoperability", in "CUDA Driver API", TRM-06703-001, Version 5.5,
3218 for additional information on library interoperability.}.
3219 This chapter describes the use cases and what changes are
3220 required in order to use both the OpenACC library and the CUBLAS and Runtime
3221 libraries within a program.
3223 @section First invocation: NVIDIA CUBLAS library API
3225 In this first use case (see below), a function in the CUBLAS library is called
3226 prior to any of the functions in the OpenACC library. More specifically, the
3227 function @code{cublasCreate()}.
3229 When invoked, the function initializes the library and allocates the
3230 hardware resources on the host and the device on behalf of the caller. Once
3231 the initialization and allocation has completed, a handle is returned to the
3232 caller. The OpenACC library also requires initialization and allocation of
3233 hardware resources. Since the CUBLAS library has already allocated the
3234 hardware resources for the device, all that is left to do is to initialize
3235 the OpenACC library and acquire the hardware resources on the host.
3237 Prior to calling the OpenACC function that initializes the library and
3238 allocate the host hardware resources, you need to acquire the device number
3239 that was allocated during the call to @code{cublasCreate()}. The invoking of the
3240 runtime library function @code{cudaGetDevice()} accomplishes this. Once
3241 acquired, the device number is passed along with the device type as
3242 parameters to the OpenACC library function @code{acc_set_device_num()}.
3244 Once the call to @code{acc_set_device_num()} has completed, the OpenACC
3245 library uses the context that was created during the call to
3246 @code{cublasCreate()}. In other words, both libraries will be sharing the
3250 /* Create the handle */
3251 s = cublasCreate(&h);
3252 if (s != CUBLAS_STATUS_SUCCESS)
3254 fprintf(stderr, "cublasCreate failed %d\n", s);
3258 /* Get the device number */
3259 e = cudaGetDevice(&dev);
3260 if (e != cudaSuccess)
3262 fprintf(stderr, "cudaGetDevice failed %d\n", e);
3266 /* Initialize OpenACC library and use device 'dev' */
3267 acc_set_device_num(dev, acc_device_nvidia);
3272 @section First invocation: OpenACC library API
3274 In this second use case (see below), a function in the OpenACC library is
3275 called prior to any of the functions in the CUBLAS library. More specificially,
3276 the function @code{acc_set_device_num()}.
3278 In the use case presented here, the function @code{acc_set_device_num()}
3279 is used to both initialize the OpenACC library and allocate the hardware
3280 resources on the host and the device. In the call to the function, the
3281 call parameters specify which device to use and what device
3282 type to use, i.e., @code{acc_device_nvidia}. It should be noted that this
3283 is but one method to initialize the OpenACC library and allocate the
3284 appropriate hardware resources. Other methods are available through the
3285 use of environment variables and these will be discussed in the next section.
3287 Once the call to @code{acc_set_device_num()} has completed, other OpenACC
3288 functions can be called as seen with multiple calls being made to
3289 @code{acc_copyin()}. In addition, calls can be made to functions in the
3290 CUBLAS library. In the use case a call to @code{cublasCreate()} is made
3291 subsequent to the calls to @code{acc_copyin()}.
3292 As seen in the previous use case, a call to @code{cublasCreate()}
3293 initializes the CUBLAS library and allocates the hardware resources on the
3294 host and the device. However, since the device has already been allocated,
3295 @code{cublasCreate()} will only initialize the CUBLAS library and allocate
3296 the appropriate hardware resources on the host. The context that was created
3297 as part of the OpenACC initialization is shared with the CUBLAS library,
3298 similarly to the first use case.
3303 acc_set_device_num(dev, acc_device_nvidia);
3305 /* Copy the first set to the device */
3306 d_X = acc_copyin(&h_X[0], N * sizeof (float));
3309 fprintf(stderr, "copyin error h_X\n");
3313 /* Copy the second set to the device */
3314 d_Y = acc_copyin(&h_Y1[0], N * sizeof (float));
3317 fprintf(stderr, "copyin error h_Y1\n");
3321 /* Create the handle */
3322 s = cublasCreate(&h);
3323 if (s != CUBLAS_STATUS_SUCCESS)
3325 fprintf(stderr, "cublasCreate failed %d\n", s);
3329 /* Perform saxpy using CUBLAS library function */
3330 s = cublasSaxpy(h, N, &alpha, d_X, 1, d_Y, 1);
3331 if (s != CUBLAS_STATUS_SUCCESS)
3333 fprintf(stderr, "cublasSaxpy failed %d\n", s);
3337 /* Copy the results from the device */
3338 acc_memcpy_from_device(&h_Y1[0], d_Y, N * sizeof (float));
3343 @section OpenACC library and environment variables
3345 There are two environment variables associated with the OpenACC library
3346 that may be used to control the device type and device number:
3347 @env{ACC_DEVICE_TYPE} and @env{ACC_DEVICE_NUM}, respectively. These two
3348 environment variables can be used as an alternative to calling
3349 @code{acc_set_device_num()}. As seen in the second use case, the device
3350 type and device number were specified using @code{acc_set_device_num()}.
3351 If however, the aforementioned environment variables were set, then the
3352 call to @code{acc_set_device_num()} would not be required.
3355 The use of the environment variables is only relevant when an OpenACC function
3356 is called prior to a call to @code{cudaCreate()}. If @code{cudaCreate()}
3357 is called prior to a call to an OpenACC function, then you must call
3358 @code{acc_set_device_num()}@footnote{More complete information
3359 about @env{ACC_DEVICE_TYPE} and @env{ACC_DEVICE_NUM} can be found in
3360 sections 4.1 and 4.2 of the @uref{https://www.openacc.org, OpenACC}
3361 Application Programming Interface”, Version 2.6.}
3365 @c ---------------------------------------------------------------------
3366 @c OpenACC Profiling Interface
3367 @c ---------------------------------------------------------------------
3369 @node OpenACC Profiling Interface
3370 @chapter OpenACC Profiling Interface
3372 @section Implementation Status and Implementation-Defined Behavior
3374 We're implementing the OpenACC Profiling Interface as defined by the
3375 OpenACC 2.6 specification. We're clarifying some aspects here as
3376 @emph{implementation-defined behavior}, while they're still under
3377 discussion within the OpenACC Technical Committee.
3379 This implementation is tuned to keep the performance impact as low as
3380 possible for the (very common) case that the Profiling Interface is
3381 not enabled. This is relevant, as the Profiling Interface affects all
3382 the @emph{hot} code paths (in the target code, not in the offloaded
3383 code). Users of the OpenACC Profiling Interface can be expected to
3384 understand that performance will be impacted to some degree once the
3385 Profiling Interface has gotten enabled: for example, because of the
3386 @emph{runtime} (libgomp) calling into a third-party @emph{library} for
3387 every event that has been registered.
3389 We're not yet accounting for the fact that @cite{OpenACC events may
3390 occur during event processing}.
3391 We just handle one case specially, as required by CUDA 9.0
3392 @command{nvprof}, that @code{acc_get_device_type}
3393 (@ref{acc_get_device_type})) may be called from
3394 @code{acc_ev_device_init_start}, @code{acc_ev_device_init_end}
3397 We're not yet implementing initialization via a
3398 @code{acc_register_library} function that is either statically linked
3399 in, or dynamically via @env{LD_PRELOAD}.
3400 Initialization via @code{acc_register_library} functions dynamically
3401 loaded via the @env{ACC_PROFLIB} environment variable does work, as
3402 does directly calling @code{acc_prof_register},
3403 @code{acc_prof_unregister}, @code{acc_prof_lookup}.
3405 As currently there are no inquiry functions defined, calls to
3406 @code{acc_prof_lookup} will always return @code{NULL}.
3408 There aren't separate @emph{start}, @emph{stop} events defined for the
3409 event types @code{acc_ev_create}, @code{acc_ev_delete},
3410 @code{acc_ev_alloc}, @code{acc_ev_free}. It's not clear if these
3411 should be triggered before or after the actual device-specific call is
3412 made. We trigger them after.
3414 Remarks about data provided to callbacks:
3418 @item @code{acc_prof_info.event_type}
3419 It's not clear if for @emph{nested} event callbacks (for example,
3420 @code{acc_ev_enqueue_launch_start} as part of a parent compute
3421 construct), this should be set for the nested event
3422 (@code{acc_ev_enqueue_launch_start}), or if the value of the parent
3423 construct should remain (@code{acc_ev_compute_construct_start}). In
3424 this implementation, the value will generally correspond to the
3425 innermost nested event type.
3427 @item @code{acc_prof_info.device_type}
3431 For @code{acc_ev_compute_construct_start}, and in presence of an
3432 @code{if} clause with @emph{false} argument, this will still refer to
3433 the offloading device type.
3434 It's not clear if that's the expected behavior.
3437 Complementary to the item before, for
3438 @code{acc_ev_compute_construct_end}, this is set to
3439 @code{acc_device_host} in presence of an @code{if} clause with
3440 @emph{false} argument.
3441 It's not clear if that's the expected behavior.
3445 @item @code{acc_prof_info.thread_id}
3446 Always @code{-1}; not yet implemented.
3448 @item @code{acc_prof_info.async}
3452 Not yet implemented correctly for
3453 @code{acc_ev_compute_construct_start}.
3456 In a compute construct, for host-fallback
3457 execution/@code{acc_device_host} it will always be
3458 @code{acc_async_sync}.
3459 It's not clear if that's the expected behavior.
3462 For @code{acc_ev_device_init_start} and @code{acc_ev_device_init_end},
3463 it will always be @code{acc_async_sync}.
3464 It's not clear if that's the expected behavior.
3468 @item @code{acc_prof_info.async_queue}
3469 There is no @cite{limited number of asynchronous queues} in libgomp.
3470 This will always have the same value as @code{acc_prof_info.async}.
3472 @item @code{acc_prof_info.src_file}
3473 Always @code{NULL}; not yet implemented.
3475 @item @code{acc_prof_info.func_name}
3476 Always @code{NULL}; not yet implemented.
3478 @item @code{acc_prof_info.line_no}
3479 Always @code{-1}; not yet implemented.
3481 @item @code{acc_prof_info.end_line_no}
3482 Always @code{-1}; not yet implemented.
3484 @item @code{acc_prof_info.func_line_no}
3485 Always @code{-1}; not yet implemented.
3487 @item @code{acc_prof_info.func_end_line_no}
3488 Always @code{-1}; not yet implemented.
3490 @item @code{acc_event_info.event_type}, @code{acc_event_info.*.event_type}
3491 Relating to @code{acc_prof_info.event_type} discussed above, in this
3492 implementation, this will always be the same value as
3493 @code{acc_prof_info.event_type}.
3495 @item @code{acc_event_info.*.parent_construct}
3499 Will be @code{acc_construct_parallel} for all OpenACC compute
3500 constructs as well as many OpenACC Runtime API calls; should be the
3501 one matching the actual construct, or
3502 @code{acc_construct_runtime_api}, respectively.
3505 Will be @code{acc_construct_enter_data} or
3506 @code{acc_construct_exit_data} when processing variable mappings
3507 specified in OpenACC @emph{declare} directives; should be
3508 @code{acc_construct_declare}.
3511 For implicit @code{acc_ev_device_init_start},
3512 @code{acc_ev_device_init_end}, and explicit as well as implicit
3513 @code{acc_ev_alloc}, @code{acc_ev_free},
3514 @code{acc_ev_enqueue_upload_start}, @code{acc_ev_enqueue_upload_end},
3515 @code{acc_ev_enqueue_download_start}, and
3516 @code{acc_ev_enqueue_download_end}, will be
3517 @code{acc_construct_parallel}; should reflect the real parent
3522 @item @code{acc_event_info.*.implicit}
3523 For @code{acc_ev_alloc}, @code{acc_ev_free},
3524 @code{acc_ev_enqueue_upload_start}, @code{acc_ev_enqueue_upload_end},
3525 @code{acc_ev_enqueue_download_start}, and
3526 @code{acc_ev_enqueue_download_end}, this currently will be @code{1}
3527 also for explicit usage.
3529 @item @code{acc_event_info.data_event.var_name}
3530 Always @code{NULL}; not yet implemented.
3532 @item @code{acc_event_info.data_event.host_ptr}
3533 For @code{acc_ev_alloc}, and @code{acc_ev_free}, this is always
3536 @item @code{typedef union acc_api_info}
3537 @dots{} as printed in @cite{5.2.3. Third Argument: API-Specific
3538 Information}. This should obviously be @code{typedef @emph{struct}
3541 @item @code{acc_api_info.device_api}
3542 Possibly not yet implemented correctly for
3543 @code{acc_ev_compute_construct_start},
3544 @code{acc_ev_device_init_start}, @code{acc_ev_device_init_end}:
3545 will always be @code{acc_device_api_none} for these event types.
3546 For @code{acc_ev_enter_data_start}, it will be
3547 @code{acc_device_api_none} in some cases.
3549 @item @code{acc_api_info.device_type}
3550 Always the same as @code{acc_prof_info.device_type}.
3552 @item @code{acc_api_info.vendor}
3553 Always @code{-1}; not yet implemented.
3555 @item @code{acc_api_info.device_handle}
3556 Always @code{NULL}; not yet implemented.
3558 @item @code{acc_api_info.context_handle}
3559 Always @code{NULL}; not yet implemented.
3561 @item @code{acc_api_info.async_handle}
3562 Always @code{NULL}; not yet implemented.
3566 Remarks about certain event types:
3570 @item @code{acc_ev_device_init_start}, @code{acc_ev_device_init_end}
3574 @c See 'DEVICE_INIT_INSIDE_COMPUTE_CONSTRUCT' in
3575 @c 'libgomp.oacc-c-c++-common/acc_prof-kernels-1.c',
3576 @c 'libgomp.oacc-c-c++-common/acc_prof-parallel-1.c'.
3577 Whan a compute construct triggers implicit
3578 @code{acc_ev_device_init_start} and @code{acc_ev_device_init_end}
3579 events, they currently aren't @emph{nested within} the corresponding
3580 @code{acc_ev_compute_construct_start} and
3581 @code{acc_ev_compute_construct_end}, but they're currently observed
3582 @emph{before} @code{acc_ev_compute_construct_start}.
3583 It's not clear what to do: the standard asks us provide a lot of
3584 details to the @code{acc_ev_compute_construct_start} callback, without
3585 (implicitly) initializing a device before?
3588 Callbacks for these event types will not be invoked for calls to the
3589 @code{acc_set_device_type} and @code{acc_set_device_num} functions.
3590 It's not clear if they should be.
3594 @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}
3598 Callbacks for these event types will also be invoked for OpenACC
3599 @emph{host_data} constructs.
3600 It's not clear if they should be.
3603 Callbacks for these event types will also be invoked when processing
3604 variable mappings specified in OpenACC @emph{declare} directives.
3605 It's not clear if they should be.
3611 Callbacks for the following event types will be invoked, but dispatch
3612 and information provided therein has not yet been thoroughly reviewed:
3615 @item @code{acc_ev_alloc}
3616 @item @code{acc_ev_free}
3617 @item @code{acc_ev_update_start}, @code{acc_ev_update_end}
3618 @item @code{acc_ev_enqueue_upload_start}, @code{acc_ev_enqueue_upload_end}
3619 @item @code{acc_ev_enqueue_download_start}, @code{acc_ev_enqueue_download_end}
3622 During device initialization, and finalization, respectively,
3623 callbacks for the following event types will not yet be invoked:
3626 @item @code{acc_ev_alloc}
3627 @item @code{acc_ev_free}
3630 Callbacks for the following event types have not yet been implemented,
3631 so currently won't be invoked:
3634 @item @code{acc_ev_device_shutdown_start}, @code{acc_ev_device_shutdown_end}
3635 @item @code{acc_ev_runtime_shutdown}
3636 @item @code{acc_ev_create}, @code{acc_ev_delete}
3637 @item @code{acc_ev_wait_start}, @code{acc_ev_wait_end}
3640 For the following runtime library functions, not all expected
3641 callbacks will be invoked (mostly concerning implicit device
3645 @item @code{acc_get_num_devices}
3646 @item @code{acc_set_device_type}
3647 @item @code{acc_get_device_type}
3648 @item @code{acc_set_device_num}
3649 @item @code{acc_get_device_num}
3650 @item @code{acc_init}
3651 @item @code{acc_shutdown}
3654 Aside from implicit device initialization, for the following runtime
3655 library functions, no callbacks will be invoked for shared-memory
3656 offloading devices (it's not clear if they should be):
3659 @item @code{acc_malloc}
3660 @item @code{acc_free}
3661 @item @code{acc_copyin}, @code{acc_present_or_copyin}, @code{acc_copyin_async}
3662 @item @code{acc_create}, @code{acc_present_or_create}, @code{acc_create_async}
3663 @item @code{acc_copyout}, @code{acc_copyout_async}, @code{acc_copyout_finalize}, @code{acc_copyout_finalize_async}
3664 @item @code{acc_delete}, @code{acc_delete_async}, @code{acc_delete_finalize}, @code{acc_delete_finalize_async}
3665 @item @code{acc_update_device}, @code{acc_update_device_async}
3666 @item @code{acc_update_self}, @code{acc_update_self_async}
3667 @item @code{acc_map_data}, @code{acc_unmap_data}
3668 @item @code{acc_memcpy_to_device}, @code{acc_memcpy_to_device_async}
3669 @item @code{acc_memcpy_from_device}, @code{acc_memcpy_from_device_async}
3674 @c ---------------------------------------------------------------------
3676 @c ---------------------------------------------------------------------
3678 @node The libgomp ABI
3679 @chapter The libgomp ABI
3681 The following sections present notes on the external ABI as
3682 presented by libgomp. Only maintainers should need them.
3685 * Implementing MASTER construct::
3686 * Implementing CRITICAL construct::
3687 * Implementing ATOMIC construct::
3688 * Implementing FLUSH construct::
3689 * Implementing BARRIER construct::
3690 * Implementing THREADPRIVATE construct::
3691 * Implementing PRIVATE clause::
3692 * Implementing FIRSTPRIVATE LASTPRIVATE COPYIN and COPYPRIVATE clauses::
3693 * Implementing REDUCTION clause::
3694 * Implementing PARALLEL construct::
3695 * Implementing FOR construct::
3696 * Implementing ORDERED construct::
3697 * Implementing SECTIONS construct::
3698 * Implementing SINGLE construct::
3699 * Implementing OpenACC's PARALLEL construct::
3703 @node Implementing MASTER construct
3704 @section Implementing MASTER construct
3707 if (omp_get_thread_num () == 0)
3711 Alternately, we generate two copies of the parallel subfunction
3712 and only include this in the version run by the master thread.
3713 Surely this is not worthwhile though...
3717 @node Implementing CRITICAL construct
3718 @section Implementing CRITICAL construct
3720 Without a specified name,
3723 void GOMP_critical_start (void);
3724 void GOMP_critical_end (void);
3727 so that we don't get COPY relocations from libgomp to the main
3730 With a specified name, use omp_set_lock and omp_unset_lock with
3731 name being transformed into a variable declared like
3734 omp_lock_t gomp_critical_user_<name> __attribute__((common))
3737 Ideally the ABI would specify that all zero is a valid unlocked
3738 state, and so we wouldn't need to initialize this at
3743 @node Implementing ATOMIC construct
3744 @section Implementing ATOMIC construct
3746 The target should implement the @code{__sync} builtins.
3748 Failing that we could add
3751 void GOMP_atomic_enter (void)
3752 void GOMP_atomic_exit (void)
3755 which reuses the regular lock code, but with yet another lock
3756 object private to the library.
3760 @node Implementing FLUSH construct
3761 @section Implementing FLUSH construct
3763 Expands to the @code{__sync_synchronize} builtin.
3767 @node Implementing BARRIER construct
3768 @section Implementing BARRIER construct
3771 void GOMP_barrier (void)
3775 @node Implementing THREADPRIVATE construct
3776 @section Implementing THREADPRIVATE construct
3778 In _most_ cases we can map this directly to @code{__thread}. Except
3779 that OMP allows constructors for C++ objects. We can either
3780 refuse to support this (how often is it used?) or we can
3781 implement something akin to .ctors.
3783 Even more ideally, this ctor feature is handled by extensions
3784 to the main pthreads library. Failing that, we can have a set
3785 of entry points to register ctor functions to be called.
3789 @node Implementing PRIVATE clause
3790 @section Implementing PRIVATE clause
3792 In association with a PARALLEL, or within the lexical extent
3793 of a PARALLEL block, the variable becomes a local variable in
3794 the parallel subfunction.
3796 In association with FOR or SECTIONS blocks, create a new
3797 automatic variable within the current function. This preserves
3798 the semantic of new variable creation.
3802 @node Implementing FIRSTPRIVATE LASTPRIVATE COPYIN and COPYPRIVATE clauses
3803 @section Implementing FIRSTPRIVATE LASTPRIVATE COPYIN and COPYPRIVATE clauses
3805 This seems simple enough for PARALLEL blocks. Create a private
3806 struct for communicating between the parent and subfunction.
3807 In the parent, copy in values for scalar and "small" structs;
3808 copy in addresses for others TREE_ADDRESSABLE types. In the
3809 subfunction, copy the value into the local variable.
3811 It is not clear what to do with bare FOR or SECTION blocks.
3812 The only thing I can figure is that we do something like:
3815 #pragma omp for firstprivate(x) lastprivate(y)
3816 for (int i = 0; i < n; ++i)
3833 where the "x=x" and "y=y" assignments actually have different
3834 uids for the two variables, i.e. not something you could write
3835 directly in C. Presumably this only makes sense if the "outer"
3836 x and y are global variables.
3838 COPYPRIVATE would work the same way, except the structure
3839 broadcast would have to happen via SINGLE machinery instead.
3843 @node Implementing REDUCTION clause
3844 @section Implementing REDUCTION clause
3846 The private struct mentioned in the previous section should have
3847 a pointer to an array of the type of the variable, indexed by the
3848 thread's @var{team_id}. The thread stores its final value into the
3849 array, and after the barrier, the master thread iterates over the
3850 array to collect the values.
3853 @node Implementing PARALLEL construct
3854 @section Implementing PARALLEL construct
3857 #pragma omp parallel
3866 void subfunction (void *data)
3873 GOMP_parallel_start (subfunction, &data, num_threads);
3874 subfunction (&data);
3875 GOMP_parallel_end ();
3879 void GOMP_parallel_start (void (*fn)(void *), void *data, unsigned num_threads)
3882 The @var{FN} argument is the subfunction to be run in parallel.
3884 The @var{DATA} argument is a pointer to a structure used to
3885 communicate data in and out of the subfunction, as discussed
3886 above with respect to FIRSTPRIVATE et al.
3888 The @var{NUM_THREADS} argument is 1 if an IF clause is present
3889 and false, or the value of the NUM_THREADS clause, if
3892 The function needs to create the appropriate number of
3893 threads and/or launch them from the dock. It needs to
3894 create the team structure and assign team ids.
3897 void GOMP_parallel_end (void)
3900 Tears down the team and returns us to the previous @code{omp_in_parallel()} state.
3904 @node Implementing FOR construct
3905 @section Implementing FOR construct
3908 #pragma omp parallel for
3909 for (i = lb; i <= ub; i++)
3916 void subfunction (void *data)
3919 while (GOMP_loop_static_next (&_s0, &_e0))
3922 for (i = _s0; i < _e1; i++)
3925 GOMP_loop_end_nowait ();
3928 GOMP_parallel_loop_static (subfunction, NULL, 0, lb, ub+1, 1, 0);
3930 GOMP_parallel_end ();
3934 #pragma omp for schedule(runtime)
3935 for (i = 0; i < n; i++)
3944 if (GOMP_loop_runtime_start (0, n, 1, &_s0, &_e0))
3947 for (i = _s0, i < _e0; i++)
3949 @} while (GOMP_loop_runtime_next (&_s0, _&e0));
3954 Note that while it looks like there is trickiness to propagating
3955 a non-constant STEP, there isn't really. We're explicitly allowed
3956 to evaluate it as many times as we want, and any variables involved
3957 should automatically be handled as PRIVATE or SHARED like any other
3958 variables. So the expression should remain evaluable in the
3959 subfunction. We can also pull it into a local variable if we like,
3960 but since its supposed to remain unchanged, we can also not if we like.
3962 If we have SCHEDULE(STATIC), and no ORDERED, then we ought to be
3963 able to get away with no work-sharing context at all, since we can
3964 simply perform the arithmetic directly in each thread to divide up
3965 the iterations. Which would mean that we wouldn't need to call any
3968 There are separate routines for handling loops with an ORDERED
3969 clause. Bookkeeping for that is non-trivial...
3973 @node Implementing ORDERED construct
3974 @section Implementing ORDERED construct
3977 void GOMP_ordered_start (void)
3978 void GOMP_ordered_end (void)
3983 @node Implementing SECTIONS construct
3984 @section Implementing SECTIONS construct
3989 #pragma omp sections
4003 for (i = GOMP_sections_start (3); i != 0; i = GOMP_sections_next ())
4020 @node Implementing SINGLE construct
4021 @section Implementing SINGLE construct
4035 if (GOMP_single_start ())
4043 #pragma omp single copyprivate(x)
4050 datap = GOMP_single_copy_start ();
4055 GOMP_single_copy_end (&data);
4064 @node Implementing OpenACC's PARALLEL construct
4065 @section Implementing OpenACC's PARALLEL construct
4068 void GOACC_parallel ()
4073 @c ---------------------------------------------------------------------
4075 @c ---------------------------------------------------------------------
4077 @node Reporting Bugs
4078 @chapter Reporting Bugs
4080 Bugs in the GNU Offloading and Multi Processing Runtime Library should
4081 be reported via @uref{https://gcc.gnu.org/bugzilla/, Bugzilla}. Please add
4082 "openacc", or "openmp", or both to the keywords field in the bug
4083 report, as appropriate.
4087 @c ---------------------------------------------------------------------
4088 @c GNU General Public License
4089 @c ---------------------------------------------------------------------
4091 @include gpl_v3.texi
4095 @c ---------------------------------------------------------------------
4096 @c GNU Free Documentation License
4097 @c ---------------------------------------------------------------------
4103 @c ---------------------------------------------------------------------
4104 @c Funding Free Software
4105 @c ---------------------------------------------------------------------
4107 @include funding.texi
4109 @c ---------------------------------------------------------------------
4111 @c ---------------------------------------------------------------------
4114 @unnumbered Library Index