Launch API to support the triple-chevron syntax

Launch API to support the triple-chevron syntax#

HIP Runtime API Reference: Launch API to support the triple-chevron syntax
Launch API to support the triple-chevron syntax
Collaboration diagram for Launch API to support the triple-chevron syntax:

Functions

hipError_t hipConfigureCall (dim3 gridDim, dim3 blockDim, size_t sharedMem, hipStream_t stream)
 Configure a kernel launch.
 
hipError_t hipSetupArgument (const void *arg, size_t size, size_t offset)
 Set a kernel argument.
 
hipError_t hipLaunchByPtr (const void *func)
 Launch a kernel.
 
hipError_t __hipPushCallConfiguration (dim3 gridDim, dim3 blockDim, size_t sharedMem, hipStream_t stream)
 Push configuration of a kernel launch.
 
hipError_t __hipPopCallConfiguration (dim3 *gridDim, dim3 *blockDim, size_t *sharedMem, hipStream_t *stream)
 Pop configuration of a kernel launch.
 
hipError_t hipLaunchKernel (const void *function_address, dim3 numBlocks, dim3 dimBlocks, void **args, size_t sharedMemBytes, hipStream_t stream)
 C compliant kernel launch API.
 
hipError_t hipLaunchHostFunc (hipStream_t stream, hipHostFn_t fn, void *userData)
 Enqueues a host function call in a stream.
 
hipError_t hipDrvMemcpy2DUnaligned (const hip_Memcpy2D *pCopy)
 
hipError_t hipExtLaunchKernel (const void *function_address, dim3 numBlocks, dim3 dimBlocks, void **args, size_t sharedMemBytes, hipStream_t stream, hipEvent_t startEvent, hipEvent_t stopEvent, int flags)
 Launches kernel from the pointer address, with arguments and shared memory on stream.
 

Detailed Description



This section describes the API to support the triple-chevron syntax.

Function Documentation

◆ __hipPopCallConfiguration()

hipError_t __hipPopCallConfiguration ( dim3 gridDim,
dim3 blockDim,
size_t *  sharedMem,
hipStream_t stream 
)

Pop configuration of a kernel launch.

Parameters
[out]gridDimgrid dimension specified as multiple of blockDim.
[out]blockDimblock dimensions specified in work-items
[out]sharedMemAmount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.
[out]streamStream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.

Please note, HIP does not support kernel launch with total work items defined in dimension with size gridDim x blockDim >= 2^32.

Please note, HIP does not support kernel launch with total work items defined in dimension with size gridDim x blockDim >= 2^32.

Returns
hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue

◆ __hipPushCallConfiguration()

hipError_t __hipPushCallConfiguration ( dim3  gridDim,
dim3  blockDim,
size_t  sharedMem,
hipStream_t  stream 
)

Push configuration of a kernel launch.

Parameters
[in]gridDimgrid dimension specified as multiple of blockDim.
[in]blockDimblock dimensions specified in work-items
[in]sharedMemAmount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.
[in]streamStream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.

Please note, HIP does not support kernel launch with total work items defined in dimension with size gridDim x blockDim >= 2^32.

Returns
hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue

◆ hipConfigureCall()

hipError_t hipConfigureCall ( dim3  gridDim,
dim3  blockDim,
size_t  sharedMem,
hipStream_t  stream 
)

Configure a kernel launch.

Parameters
[in]gridDimgrid dimension specified as multiple of blockDim.
[in]blockDimblock dimensions specified in work-items
[in]sharedMemAmount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.
[in]streamStream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.

Please note, HIP does not support kernel launch with total work items defined in dimension with size gridDim x blockDim >= 2^32.

Returns
hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue

◆ hipDrvMemcpy2DUnaligned()

hipError_t hipDrvMemcpy2DUnaligned ( const hip_Memcpy2D pCopy)

Copies memory for 2D arrays.

Parameters
pCopy- Parameters for the memory copy
Returns
hipSuccess, hipErrorInvalidValue

◆ hipExtLaunchKernel()

hipError_t hipExtLaunchKernel ( const void *  function_address,
dim3  numBlocks,
dim3  dimBlocks,
void **  args,
size_t  sharedMemBytes,
hipStream_t  stream,
hipEvent_t  startEvent,
hipEvent_t  stopEvent,
int  flags 
)

Launches kernel from the pointer address, with arguments and shared memory on stream.

Parameters
[in]function_addresspointer to the Kernel to launch.
[in]numBlocksnumber of blocks.
[in]dimBlocksdimension of a block.
[in]argspointer to kernel arguments.
[in]sharedMemBytesAmount of dynamic shared memory to allocate for this kernel. HIP-Clang compiler provides support for extern shared declarations.
[in]streamStream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.
[in]startEventIf non-null, specified event will be updated to track the start time of the kernel launch. The event must be created before calling this API.
[in]stopEventIf non-null, specified event will be updated to track the stop time of the kernel launch. The event must be created before calling this API.
[in]flagsThe value of hipExtAnyOrderLaunch, signifies if kernel can be launched in any order.
Returns
hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue.

◆ hipLaunchByPtr()

hipError_t hipLaunchByPtr ( const void *  func)

Launch a kernel.

Parameters
[in]funcKernel to launch.
Returns
hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue

◆ hipLaunchHostFunc()

hipError_t hipLaunchHostFunc ( hipStream_t  stream,
hipHostFn_t  fn,
void *  userData 
)

Enqueues a host function call in a stream.

Parameters
[in]stream- The stream to enqueue work in.
[in]fn- The function to call once enqueued preceeding operations are complete.
[in]userData- User-specified data to be passed to the function.
Returns
hipSuccess, hipErrorInvalidResourceHandle, hipErrorInvalidValue, hipErrorNotSupported

The host function to call in this API will be executed after the preceding operations in the stream are complete. The function is a blocking operation that blocks operations in the stream that follow it, until the function is returned. Event synchronization and internal callback functions make sure enqueued operations will execute in order, in the stream.

The host function must not make any HIP API calls. The host function is non-reentrant. It must not perform sychronization with any operation that may depend on other processing execution but is not enqueued to run earlier in the stream.

Host functions that are enqueued respectively in different non-blocking streams can run concurrently.

Warning
This API is marked as beta, meaning, while this is feature complete, it is still open to changes and may have outstanding issues.

◆ hipLaunchKernel()

hipError_t hipLaunchKernel ( const void *  function_address,
dim3  numBlocks,
dim3  dimBlocks,
void **  args,
size_t  sharedMemBytes,
hipStream_t  stream 
)

C compliant kernel launch API.

Parameters
[in]function_address- kernel stub function pointer.
[in]numBlocks- number of blocks
[in]dimBlocks- dimension of a block
[in]args- kernel arguments
[in]sharedMemBytes- Amount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.
[in]stream- Stream where the kernel should be dispatched. May be 0, in which case th default stream is used with associated synchronization rules.
Returns
hipSuccess, hipErrorInvalidValue

◆ hipSetupArgument()

hipError_t hipSetupArgument ( const void *  arg,
size_t  size,
size_t  offset 
)

Set a kernel argument.

Returns
hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue
Parameters
[in]argPointer the argument in host memory.
[in]sizeSize of the argument.
[in]offsetOffset of the argument on the argument stack.