cubecl_hip_sys

Function hipModuleLaunchKernel

Source
pub unsafe extern "C" fn hipModuleLaunchKernel(
    f: hipFunction_t,
    gridDimX: c_uint,
    gridDimY: c_uint,
    gridDimZ: c_uint,
    blockDimX: c_uint,
    blockDimY: c_uint,
    blockDimZ: c_uint,
    sharedMemBytes: c_uint,
    stream: hipStream_t,
    kernelParams: *mut *mut c_void,
    extra: *mut *mut c_void,
) -> hipError_t
Expand description

@brief launches kernel f with launch parameters and shared memory on stream with arguments passed to kernelparams or extra

@param [in] f Kernel to launch. @param [in] gridDimX X grid dimension specified as multiple of blockDimX. @param [in] gridDimY Y grid dimension specified as multiple of blockDimY. @param [in] gridDimZ Z grid dimension specified as multiple of blockDimZ. @param [in] blockDimX X block dimensions specified in work-items @param [in] blockDimY Y grid dimension specified in work-items @param [in] blockDimZ Z grid dimension specified in work-items @param [in] sharedMemBytes Amount of dynamic shared memory to allocate for this kernel. The HIP-Clang compiler provides support for extern shared declarations. @param [in] stream Stream where the kernel should be dispatched. May be 0, in which case th default stream is used with associated synchronization rules. @param [in] kernelParams Kernel parameters to launch @param [in] extra Pointer to kernel arguments. These are passed directly to the kernel and must be in the memory layout and alignment expected by the kernel. All passed arguments must be naturally aligned according to their type. The memory address of each argument should be a multiple of its size in bytes. Please refer to hip_porting_driver_api.md for sample usage.

Please note, HIP does not support kernel launch with total work items defined in dimension with size gridDim x blockDim >= 2^32. So gridDim.x * blockDim.x, gridDim.y * blockDim.y and gridDim.z * blockDim.z are always less than 2^32.

@returns #hipSuccess, #hipErrorNotInitialized, #hipErrorInvalidValue