MPIX_Win_complete_stream

Completes an RMA access epoch with GPU stream awareness

Definitions

C/C++ Synopsis

int MPIX_Win_complete_stream ( MPI_Win win, void * stream )

Arguments

win                 window object (handle) passed as input
stream              GPU stream handle passed as input

Description

MPIX_Win_complete_stream enqueues operations to GPU stream stream to complete an RMA access epoch on win created by an MPI_Win_start operation with MPIX_MODE_STREAM. All access epochs created with stream awareness are closed using MPIX_Win_complete_stream.

MPIX_Win_complete_stream enqueues operations into the GPU stream stream to enforce the completion of all preceding GPU stream aware RMA calls at the origin. No remote visibility guarantees are provided.

MPIX_Win_complete_stream is a non-blocking operation with respect to the application process. Necessary trigger, completion, and signaling events are enqueued into the execution stream to enforce the completion of all preceding GPU stream aware RMA calls at the origin and ensure closure of the active access epoch on the win.

Return Values

None.

Examples

C/C++ Example

Example code snippet showing the usage of basic stream triggered communication operations and comparing about regular non-GPU stream aware MPI active RMA operation.

Non-GPU Stream-aware Example

for (int i = 0; i < iterations; i++ ) {
    MPI_Win_post(group,0,win);

    increment<<<nb,nt,0,stream>>>(n,source);
    CHECK(hipStreamSynchronize(stream));

    MPI_Win_start(group,0,win);
    MPI_Put(source,n,MPI_INT,left,fromL-base,n,MPI_INT,win);
    MPI_Put(source,n,MPI_INT,right,fromR-base,n,MPI_INT,win);
    MPI_Win_complete(win);

    MPI_Win_wait(win);

    compare<<<nb,nt,0,stream>>>(i,n,fromL);
    compare<<<nb,nt,0,stream>>>(i,n,fromR);
    CHECK(hipStreamSynchronize(stream));
}

GPU Stream-aware Example

for (int i = 0; i < iterations; i++ ) {
    MPIX_Win_post_stream(group,win,stream);
    increment<<<nb,nt,0,stream>>>(n,source);

    MPI_Win_start(group,MPI_MODE_STREAM,win);
    MPI_Put(source,n,MPI_INT,left,fromL-base,n,MPI_INT,win);
    MPI_Put(source,n,MPI_INT,right,fromR-base,n,MPI_INT,win);
    MPIX_Win_complete_stream(win,stream);

    MPIX_Win_wait_stream(win,stream);

    compare<<<nb,nt,0,stream>>>(i,n,fromL);
    compare<<<nb,nt,0,stream>>>(i,n,fromR);
}
CHECK(hipStreamSynchronize(stream));