Launch API

Launch API#

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

Configure a kernel launch.

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

Parameters:

  • gridDim – [in] grid dimension specified as multiple of blockDim.

  • blockDim – [in] block dimensions specified in work-items

  • sharedMem – [in] Amount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.

  • stream – [in] Stream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.

Returns:

hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue

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

Set a kernel argument.

Parameters:

  • arg – [in] Pointer the argument in host memory.

  • size – [in] Size of the argument.

  • offset – [in] Offset of the argument on the argument stack.

Returns:

hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue

hipError_t hipLaunchByPtr(const void *func)#

Launch a kernel.

Parameters:

func – [in] Kernel to launch.

Returns:

hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue

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

Push configuration of a kernel launch.

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

Parameters:

  • gridDim – [in] grid dimension specified as multiple of blockDim.

  • blockDim – [in] block dimensions specified in work-items

  • sharedMem – [in] Amount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.

  • stream – [in] Stream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.

Returns:

hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue

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

Pop configuration of a kernel launch.

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.

Parameters:

  • gridDim – [out] grid dimension specified as multiple of blockDim.

  • blockDim – [out] block dimensions specified in work-items

  • sharedMem – [out] Amount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.

  • stream – [out] Stream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.

Returns:

hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue

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:

  • function_address – [in] - kernel stub function pointer.

  • numBlocks – [in] - number of blocks

  • dimBlocks – [in] - dimension of a block

  • args – [in] - kernel arguments

  • sharedMemBytes – [in] - Amount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.

  • stream – [in] - 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

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

Enqueues a host function call in a stream.

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.

Parameters:

  • stream – [in] - The stream to enqueue work in.

  • fn – [in] - The function to call once enqueued preceeding operations are complete.

  • userData – [in] - User-specified data to be passed to the function.

Returns:

hipSuccess, hipErrorInvalidResourceHandle, hipErrorInvalidValue, hipErrorNotSupported

hipError_t hipDrvMemcpy2DUnaligned(const hip_Memcpy2D *pCopy)#

Copies memory for 2D arrays.

Parameters:

pCopy – - Parameters for the memory copy

Returns:

hipSuccess, hipErrorInvalidValue

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:

  • function_address – [in] pointer to the Kernel to launch.

  • numBlocks – [in] number of blocks.

  • dimBlocks – [in] dimension of a block.

  • args – [in] pointer to kernel arguments.

  • sharedMemBytes – [in] Amount of dynamic shared memory to allocate for this kernel. HIP-Clang compiler provides support for extern shared declarations.

  • stream – [in] Stream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.

  • startEvent – [in] If 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.

  • stopEvent – [in] If 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.

  • flags – [in] The value of hipExtAnyOrderLaunch, signifies if kernel can be launched in any order.

Returns:

hipSuccess, hipErrorNotInitialized, hipErrorInvalidValue.