Ctrl+K

Memory Ordering

Memory Ordering#

List of Operations affected by Memory Ordering Routines:

Operations

Fence

Quiet

Memory Store

X

X

Blocking Put

X

X

Blocking Get

Blocking Fetching AMO

Blocking Non-fetching AMO

X

X

Blocking put-with-signal

X

X

Nonblocking Put

X

X

Nonblocking Get

X

Nonblocking AMO

X [1]

X

Nonblocking put-with-signal

X

X

This section introduces Intel® SHMEM interfaces that provide mechanisms to ensure ordering and/or delivery of completion on memory store, blocking, and nonblocking routines. Table List of Operations affected by Memory Ordering Routines lists the operations affected by Intel® SHMEM memory ordering routines.

ISHMEM_FENCE#

Ensures ordering of delivery of operations on symmetric data objects.

void ishmem_fence(void)#
Parameters:

None.

Returns:

None.

Callable from the host and device.

Description: This routine ensures ordering of delivery of operations on symmetric data objects.

All operations on symmetric data objects issued to a particular PE prior to the call to ishmem_fence are guaranteed to be delivered before any subsequent operations on symmetric data objects to the same PE. ishmem_fence guarantees order of delivery, not completion. It does not guarantee order of delivery of nonblocking Get or values fetched by nonblocking AMO routines.

ISHMEMX_FENCE_WORK_GROUP#

Ensures ordering of delivery of operations on symmetric data objects.

void ishmemx_fence_work_group(const Group &group)#
Parameters:

group – The SYCL group or sub_group on which to collectively perform the fence operation.

Returns:

None.

Callable from the device.

Description: This routine ensures ordering of delivery of operations on symmetric data objects.

All operations on symmetric data objects issued to a particular PE prior to the call to ishmemx_fence_work_group are guaranteed to be delivered before any subsequent operations on symmetric data objects to the same PE. ishmemx_fence_work_group guarantees order of delivery, not completion. It does not guarantee order of delivery of nonblocking Get or values fetched by nonblocking AMO routines.

ISHMEM_QUIET#

Waits for completion of outstanding operations on symmetric data objects issued by a PE.

void ishmem_quiet(void)#
Parameters:

None.

Returns:

None.

Callable from the host and device.

Description: The ishmem_quiet routine ensures completion of all operations on symmetric data objects issued by the calling PE.

All operations on symmetric data objects are guaranteed to be complete and visible to all PEs when ishmem_quiet returns.

A host-initiated ishmem_quiet will only guarantee completion of device-initiated operations for which the corresponding SYCL kernel has completed execution.

ISHMEMX_QUIET_ON_QUEUE#

Waits for completion of outstanding operations on symmetric data objects issued by a PE.

sycl::event ishmemx_quiet_on_queue(sycl::queue &q, const std::vector<sycl::event> &deps)#
Parameters:

  • q – The SYCL queue on which to execute the operation. q must be mapped to the GPU tile assigned to the calling PE.

  • deps – An optional vector of SYCL events that the operation depends on.

Returns:

The SYCL event created upon submitting the operation to the SYCL runtime.

Callable from the host.

Description: The ishmemx_quiet_on_queue routine ensures completion of all operations on symmetric data objects issued by the calling PE.

All operations on symmetric data objects are guaranteed to be complete and visible to all PEs when ishmemx_quiet_on_queue returns.

To ensure the quiet operation has completed, refer to the on_queue API Completion Semantics section.

ISHMEMX_QUIET_WORK_GROUP#

Waits for completion of outstanding operations on symmetric data objects issued by a PE.

void ishmemx_quiet_work_group(const Group &group)#
Parameters:

group – The SYCL group or sub_group on which to collectively perform the quiet operation.

Returns:

None.

Callable from the device.

Description: The ishmemx_quiet_work_group routine ensures completion of all operations on symmetric data objects issued by the calling PE.

All operations on symmetric data objects are guaranteed to be complete and visible to all PEs when ishmemx_quiet_work_group returns.

Contents