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

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]	gridDim	grid dimension specified as multiple of blockDim.
[out]	blockDim	block dimensions specified in work-items
[out]	sharedMem	Amount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations.
[out]	stream	Stream 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

◆ __hipPushCallConfiguration()

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

Push configuration of a kernel launch.

Parameters

[in]	gridDim	grid dimension specified as multiple of blockDim.
[in]	blockDim	block dimensions specified in work-items
[in]	sharedMem	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 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]	gridDim	grid dimension specified as multiple of blockDim.
[in]	blockDim	block dimensions specified in work-items
[in]	sharedMem	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 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_address	- Pointer to the Kernel to launch.
[in]	numBlocks	- Number of blocks.
[in]	dimBlocks	- Dimension of a block.
[in]	args	- Pointer of arguments passed to the kernel. If the kernel has multiple parameters, 'args' should be array of pointers, each points the corresponding argument.
[in]	sharedMemBytes	- Amount of dynamic shared memory to allocate for this kernel. HIP-Clang compiler provides support for extern shared declarations.
[in]	stream	- Stream where the kernel should be dispatched. May be 0, in which case the default stream is used with associated synchronization rules.
[in]	startEvent	- 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.
[in]	stopEvent	- 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.
[in]	flags	- The 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] func Kernel 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	- Pointer of arguments passed to the kernel. If the kernel has multiple parameters, 'args' should be array of pointers, each points the corresponding argument.
[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]	arg	Pointer the argument in host memory.
[in]	size	Size of the argument.
[in]	offset	Offset of the argument on the argument stack.

Launch API to support the triple-chevron syntax

Launch API to support the triple-chevron syntax#

Functions

Detailed Description

Function Documentation

◆ __hipPopCallConfiguration()

◆ __hipPushCallConfiguration()

◆ hipConfigureCall()

◆ hipDrvMemcpy2DUnaligned()

◆ hipExtLaunchKernel()

◆ hipLaunchByPtr()

◆ hipLaunchHostFunc()

◆ hipLaunchKernel()

◆ hipSetupArgument()