diff --git a/projects/hip/include/hip/driver_types.h b/projects/hip/include/hip/driver_types.h index 49307d4d50..b212160862 100644 --- a/projects/hip/include/hip/driver_types.h +++ b/projects/hip/include/hip/driver_types.h @@ -445,6 +445,111 @@ typedef struct HIP_MEMCPY3D { size_t Height; ///< Height in bytes of 3D memory copy size_t Depth; ///< Depth in bytes of 3D memory copy } HIP_MEMCPY3D; +/** + * Specifies the type of location + */ + typedef enum hipMemLocationType { + hipMemLocationTypeInvalid = 0, + hipMemLocationTypeDevice = 1 ///< Device location, thus it's HIP device ID +} hipMemLocationType; +/** + * Specifies a memory location. + * + * To specify a gpu, set type = @p hipMemLocationTypeDevice and set id = the gpu's device ID + */ +typedef struct hipMemLocation { + hipMemLocationType type; ///< Specifies the location type, which describes the meaning of id + int id; ///< Identifier for the provided location type @p hipMemLocationType +} hipMemLocation; + +/** + * Flags to specify for copies within a batch. Used with hipMemcpyBatchAsync + */ +typedef enum hipMemcpyFlags { + hipMemcpyFlagDefault = 0x0, ///< Default flag + hipMemcpyFlagPreferOverlapWithCompute = 0x1 ///< Tries to overlap copy with compute work. +} hipMemcpyFlags; + +/** + * Flags to specify order in which source pointer is accessed by Batch memcpy + */ +typedef enum hipMemcpySrcAccessOrder { + hipMemcpySrcAccessOrderInvalid = 0x0, ///< Default Invalid. + hipMemcpySrcAccessOrderStream = 0x1, ///< Access to source pointer must be in stream order. + hipMemcpySrcAccessOrderDuringApiCall = 0x2, ///< Access to source pointer can be out of stream order and all accesses must be complete before API call returns. + hipMemcpySrcAccessOrderAny = 0x3, ///< Access to the source pointer can be out of stream order and the accesses can happen even after the API call return. + hipMemcpySrcAccessOrderMax = 0x7FFFFFFF +} hipMemcpySrcAccessOrder; + +/** + * Attributes for copies within a batch. + */ +typedef struct hipMemcpyAttributes { + hipMemcpySrcAccessOrder srcAccessOrder; ///< Source access ordering to be observed for copies with this attribute. + hipMemLocation srcLocHint; ///< Location hint for src operand. + hipMemLocation dstLocHint; ///< Location hint for destination operand. + unsigned int flags; ///< Additional Flags for copies. See hipMemcpyFlags. +} hipMemcpyAttributes; +/** + * Operand types for individual copies within a batch + */ +typedef enum hipMemcpy3DOperandType { + hipMemcpyOperandTypePointer = 0x1, ///< Mempcy operand is a valid pointer. + hipMemcpyOperandTypeArray = 0x2, ///< Memcpy operand is a valid hipArray. + hipMemcpyOperandTypeMax = 0x7FFFFFFF +} hipMemcpy3DOperandType; + +/** + * Struct representing offset into a hipArray_t in elements. + */ +typedef struct hipOffset3D { + size_t x; + size_t y; + size_t z; +} hipOffset3D; +/** + * Struct representing an operand for copy with hipMemcpy3DBatchAsync. + */ +typedef struct hipMemcpy3DOperand { + hipMemcpy3DOperandType type; + union { + struct { + void *ptr; + size_t rowLength; ///< Length of each row in elements. + size_t layerHeight; ///< Height of each layer in elements. + hipMemLocation locHint; ///< Location Hint for the operand. + } ptr; + struct { + hipArray_t array; ///< Array struct for hipMemcpyOperandTypeArray. + hipOffset3D offset; ///< Offset into array in elements. + } array; + } op; +} hipMemcy3DOperand; + +/** + * HIP 3D Batch Op + */ +typedef struct hipMemcpy3DBatchOp { + hipMemcpy3DOperand src; + hipMemcpy3DOperand dst; + hipExtent extent; + hipMemcpySrcAccessOrder srcAccessOrder; + unsigned int flags; +} hipMemcpy3DBatchOp; + +typedef struct hipMemcpy3DPeerParms +{ + hipArray_t srcArray; ///< Source memory address + hipPos srcPos; ///< Source position offset + hipPitchedPtr srcPtr; ///< Pitched source memory address + int srcDevice; ///< Source device + hipArray_t dstArray; ///< Destination memory address + hipPos dstPos; ///< Destination position offset + hipPitchedPtr dstPtr; ///< Pitched destination memory address + int dstDevice; ///< Destination device + hipExtent extent; ///< Requested memory copy size +} hipMemcpy3DPeerParms; + /** * @brief Make hipPitchedPtr * diff --git a/projects/hip/include/hip/hip_runtime_api.h b/projects/hip/include/hip/hip_runtime_api.h index e6251e03ba..5dc3147c4f 100644 --- a/projects/hip/include/hip/hip_runtime_api.h +++ b/projects/hip/include/hip/hip_runtime_api.h @@ -1140,22 +1140,7 @@ typedef enum hipMemPoolAttr */ hipMemPoolAttrUsedMemHigh = 0x8 } hipMemPoolAttr; -/** - * Specifies the type of location - */ - typedef enum hipMemLocationType { - hipMemLocationTypeInvalid = 0, - hipMemLocationTypeDevice = 1 ///< Device location, thus it's HIP device ID -} hipMemLocationType; -/** - * Specifies a memory location. - * - * To specify a gpu, set type = @p hipMemLocationTypeDevice and set id = the gpu's device ID - */ -typedef struct hipMemLocation { - hipMemLocationType type; ///< Specifies the location type, which describes the meaning of id - int id; ///< Identifier for the provided location type @p hipMemLocationType -} hipMemLocation; + /** * Specifies the memory protection flags for mapping * @@ -5645,6 +5630,60 @@ hipError_t hipDrvMemcpy3DAsync(const HIP_MEMCPY3D* pCopy, hipStream_t stream); * hipCtxSetCurrent, hipCtxPushCurrent, hipCtxSetCacheConfig, hipCtxSynchronize, hipCtxGetDevice */ hipError_t hipMemGetAddressRange(hipDeviceptr_t* pbase, size_t* psize, hipDeviceptr_t dptr); + +/** + * @brief Perform Batch of 1D copies + * + * @param [in] dsts - Array of destination pointers + * @param [in] srcs - Array of source pointers. + * @param [in] sizes - Array of sizes for memcpy operations + * @param [in] count - Size of dsts, srcs and sizes arrays + * @param [in] attrs - Array of memcpy attributes (not supported) + * @param [in] attrsIdxs - Array of indices to map attrs to copies (not supported) + * @param [in] numAttrs - Size of attrs and attrsIdxs arrays (not supported) + * @param [in] failIdx - Pointer to a location to return failure index inside the batch + * @param [in] stream - stream used to enqueue operations in. + * + * @returns #hipSuccess, #hipErrorInvalidValue + */ +hipError_t hipMemcpyBatchAsync(void **dsts, void **srcs, size_t *sizes, size_t count, + hipMemcpyAttributes *attrs, size_t *attrsIdxs, size_t numAttrs, + size_t *failIdx, hipStream_t stream __dparm(0)); + +/** + * @brief Perform Batch of 3D copies + * + * @param [in] numOps - Total number of memcpy operations. + * @param [in] opList - Array of size numOps containing the actual memcpy operations. + * @param [in] failIdx - Pointer to a location to return the index of the copy where a failure + * - was encountered. + * @param [in] flags - Flags for future use, must be zero now. + * @param [in] stream - The stream to enqueue the operations in. + * + * @returns #hipSuccess, #hipErrorInvalidValue + */ +hipError_t hipMemcpy3DBatchAsync(size_t numOps, struct hipMemcpy3DBatchOp *opList, size_t *failIdx, + unsigned long long flags, hipStream_t stream __dparm(0)); + +/** + * @brief Performs 3D memory copies between devices + * This API is asynchronous with respect to host + * + * @param [in] p - Parameters for memory copy + * + * @returns #hipSuccess, #hipErrorInvalidValue, hipErrorInvalidDevice + */ +hipError_t hipMemcpy3DPeer(hipMemcpy3DPeerParms *p); + +/** + * @brief Performs 3D memory copies between devices asynchronously + * + * @param [in] p - Parameters for memory copy + * @param [in] stream - Stream to enqueue operation in. + * + * @returns #hipSuccess, #hipErrorInvalidValue, hipErrorInvalidDevice + */ +hipError_t hipMemcpy3DPeerAsync(hipMemcpy3DPeerParms *p, hipStream_t stream __dparm(0)); // doxygen end Memory /** * @}