Skip to content

Commit 5c9cd0c

Browse files
committed
[SYCL][Doc] Add atomic_fence to ExtendedAtomics
Signed-off-by: John Pennycook <[email protected]>
1 parent 655d35b commit 5c9cd0c

File tree

1 file changed

+17
-1
lines changed

1 file changed

+17
-1
lines changed

sycl/doc/extensions/ExtendedAtomics/SYCL_INTEL_extended_atomics.asciidoc

Lines changed: 17 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -59,7 +59,7 @@ This extension is written against the SYCL 1.2.1 specification, Revision 6.
5959

6060
== Overview
6161

62-
The SYCL atomic library (+cl::sycl::atomic+) defined in SYCL 1.2.1 is based on the standard atomic libary (+std::atomic+) but has some differences. This extension introduces an alternative atomic class (+cl::sycl::intel::atomic_ref+) including additional features from {cpp}20:
62+
The SYCL atomic library (+cl::sycl::atomic+) defined in SYCL 1.2.1 is based on the standard atomic libary (+std::atomic+) but has some differences. This extension introduces a new fence function (+cl::sycl::intel::atomic_fence+) and an alternative atomic class (+cl::sycl::intel::atomic_ref+) including additional features from {cpp}20:
6363

6464
- Overloaded operators to reduce the verbosity of using atomics
6565
- Missing functions (e.g. `is_lock_free()`)
@@ -282,6 +282,18 @@ The additional member functions below are available for atomic references to poi
282282

283283
|===
284284

285+
==== Atomic Fence
286+
287+
The +atomic_fence+ function corresponds to the +std::atomic_thread_fence+ function, and performs a memory fence ordering accesses to any memory space.
288+
289+
The effects of a call to +atomic_fence+ depend on the value of the +order+ parameter:
290+
291+
- `relaxed`: No effect
292+
- `acquire`: Acquire fence
293+
- `release`: Release fence
294+
- `acq_rel`: Both an acquire fence and a release fence
295+
- `seq_cst`: A sequentially consistent acquire and release fence
296+
285297
==== Sample Header
286298

287299
[source,c++]
@@ -489,6 +501,9 @@ class atomic_ref<T*, DefaultOrder, Space> {
489501
T* operator-=(difference_type) const noexcept;
490502
491503
};
504+
505+
void atomic_fence(memory_order order):
506+
492507
} // namespace intel
493508
} // namespace sycl
494509
} // namespace cl
@@ -513,6 +528,7 @@ None.
513528
|Rev|Date|Author|Changes
514529
|1|2020-01-30|John Pennycook|*Initial public working draft*
515530
|2|2020-04-07|John Pennycook|*Rename class, remove accessor usage, adjust memory orders*
531+
|3|2020-04-09|John Pennycook|*Add atomic_fence*
516532
|========================================
517533
518534
//************************************************************************

0 commit comments

Comments
 (0)