diff --git a/projects/hip/include/hip/hip_ext.h b/projects/hip/include/hip/hip_ext.h index 2548cbe1e5..09e9cf8b95 100644 --- a/projects/hip/include/hip/hip_ext.h +++ b/projects/hip/include/hip/hip_ext.h @@ -49,6 +49,9 @@ THE SOFTWARE. * @param [in] kernelParams pointer to kernel parameters. * @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. * @param [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. * @param [in] stopEvent If non-null, specified event will be updated to track the stop time of @@ -57,8 +60,6 @@ THE SOFTWARE. * launched in any order. * @returns #hipSuccess, #hipInvalidDeviceId, #hipErrorNotInitialized, #hipErrorInvalidValue. * - * @warning kernellParams argument is not yet implemented in HIP, use extra instead. - * Please refer to hip_porting_driver_api.md for sample usage. * HIP/ROCm actually updates the start event when the associated kernel completes. * Currently, timing between startEvent and stopEvent does not include the time it takes to perform * a system scope release/cache flush - only the time it takes to issues writes to cache. diff --git a/projects/hip/include/hip/hip_runtime_api.h b/projects/hip/include/hip/hip_runtime_api.h index a1f3822366..75326d0d70 100644 --- a/projects/hip/include/hip/hip_runtime_api.h +++ b/projects/hip/include/hip/hip_runtime_api.h @@ -5042,15 +5042,15 @@ hipError_t hipModuleLoadDataEx(hipModule_t* module, const void* image, unsigned * @param [in] kernelParams * @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, hipInvalidDevice, hipErrorNotInitialized, hipErrorInvalidValue - * - * @warning kernellParams argument is not yet implemented in HIP. Please use extra instead. Please - * refer to hip_porting_driver_api.md for sample usage. */ hipError_t hipModuleLaunchKernel(hipFunction_t f, unsigned int gridDimX, unsigned int gridDimY, unsigned int gridDimZ, unsigned int blockDimX,