AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

Merge commit 'ed7fae51d884a17cd3bd96c7973d1f1f78b5f3a5' into develop

Browse Source
This commit is contained in:
Joseph Macaranas
2025-08-18 16:23:12 -04:00
parent 32634d809f ed7fae51d8
commit a22b6f3aca
1 changed files with 204 additions and 14 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/hipother/hipnv/include/hip/nvidia_detail/nvidia_hip_runtime_api.h
+204 -14
View File
@@ -28,6 +28,7 @@ THE SOFTWARE.
#include <cuda_profiler_api.h>
#include <cuda_fp16.h>
#include <driver_types.h>
#include <stdio.h>
#define CUDA_9000 9000
@@ -300,6 +301,12 @@ typedef enum cudaMemRangeAttribute hipMemRangeAttribute;
#define hipMemRangeAttributeAccessedBy cudaMemRangeAttributeAccessedBy
#define hipMemRangeAttributeLastPrefetchLocation cudaMemRangeAttributeLastPrefetchLocation
#if CUDA_VERSION >= CUDA_12000
typedef enum CUmemRangeHandleType_enum hipMemRangeHandleType;
#define hipMemRangeHandleTypeDmaBufFd CU_MEM_RANGE_HANDLE_TYPE_DMA_BUF_FD
#define hipMemRangeHandleTypeMax CU_MEM_RANGE_HANDLE_TYPE_MAX
#endif
#define hipSurfaceBoundaryMode cudaSurfaceBoundaryMode
#define hipBoundaryModeZero cudaBoundaryModeZero
#define hipBoundaryModeTrap cudaBoundaryModeTrap
@@ -865,6 +872,29 @@ static inline void hipMemcpy2DTocudaMemcpy2D(CUDA_MEMCPY2D* a, const hip_Memcpy2
a->Height = (size_t)p->Height;
}
#if CUDA_VERSION >= CUDA_12020
typedef enum cudaMemcpyFlags hipMemcpyFlags;
#define hipMemcpyFlagDefault cudaMemcpyFlagDefault
#define hipMemcpyFlagPreferOverlapWithCompute cudaMemcpyFlagPreferOverlapWithCompute
typedef enum cudaMemcpySrcAccessOrder hipMemcpySrcAccessOrder;
#define hipMemcpySrcAccessOrderInvalid cudaMemcpySrcAccessOrderInvalid
#define hipMemcpySrcAccessOrderStream cudaMemcpySrcAccessOrderStream
#define hipMemcpySrcAccessOrderDuringApiCall cudaMemcpySrcAccessOrderDuringApiCall
#define hipMemcpySrcAccessOrderAny cudaMemcpySrcAccessOrderAny
#define hipMemcpySrcAccessOrderMax cudaMemcpySrcAccessOrderMax
typedef struct cudaMemcpyAttributes hipMemcpyAttributes;
typedef enum cudaMemcpy3DOperandType hipMemcpy3DOperandType;
#define hipMemcpyOperandTypePointer cudaMemcpyOperandTypePointer
#define hipMemcpyOperandTypeArray cudaMemcpyOperandTypeArray
#define hipMemcpyOperandTypeMax cudaMemcpyOperandTypeMax
typedef struct cudaOffset3D hipOffset3D;
typedef struct cudaMemcpy3DOperand hipMemcpy3DOperand;
typedef struct cudaMemcpy3DBatchOp hipMemcpy3DBatchOp;
typedef struct cudaMemcpy3DPeerParms hipMemcpy3DPeerParms;
#endif
#define hipMemcpy3DParms cudaMemcpy3DParms
#define hipArrayDefault cudaArrayDefault
@@ -970,6 +1000,15 @@ typedef enum cudaGraphInstantiateFlags hipGraphInstantiateFlags;
#define hipGraphInstantiateFlagDeviceLaunch cudaGraphInstantiateFlagDeviceLaunch
#define hipGraphInstantiateFlagUseNodePriority cudaGraphInstantiateFlagUseNodePriority
typedef enum cudaDriverEntryPointQueryResult hipDriverEntryPointQueryResult;
#define hipDriverEntryPointSuccess cudaDriverEntryPointSuccess
#define hipDriverEntryPointSymbolNotFound cudaDriverEntryPointSymbolNotFound
#define hipDriverEntryPointVersionNotSufficent cudaDriverEntryPointVersionNotSufficent
#define hipEnableDefault cudaEnableDefault
#define hipEnableLegacyStream cudaEnableLegacyStream
#define hipEnablePerThreadDefaultStream cudaEnablePerThreadDefaultStream
inline static hipError_t hipCUDAErrorTohipError(cudaError_t cuError) {
switch (cuError) {
case cudaSuccess:
@@ -1812,6 +1851,26 @@ typedef enum cudaKernelNodeAttrID hipKernelNodeAttrID;
#define hipKernelNodeAttributeCooperative cudaKernelNodeAttributeCooperative
#define hipKernelNodeAttributePriority cudaKernelNodeAttributePriority
#if CUDA_VERSION >= CUDA_12000
typedef union cudaStreamAttrValue hipStreamAttrValue;
typedef enum cudaStreamAttrID hipStreamAttrID;
#define hipStreamAttributeAccessPolicyWindow cudaStreamAttributeAccessPolicyWindow
#define hipStreamAttributeSynchronizationPolicy cudaStreamAttributeSynchronizationPolicy
#define hipStreamAttributeMemSyncDomainMap cudaStreamAttributeMemSyncDomainMap
#define hipStreamAttributeMemSyncDomain cudaStreamAttributeMemSyncDomain
#define hipStreamAttributePriority cudaStreamAttributePriority
typedef enum cudaSynchronizationPolicy hipSynchronizationPolicy;
#define hipSyncPolicyAuto cudaSyncPolicyAuto
#define hipSyncPolicySpin cudaSyncPolicySpin
#define hipSyncPolicyYield cudaSyncPolicyYield
#define hipSyncPolicyBlockingSync cudaSyncPolicyBlockingSync
typedef enum cudaLaunchMemSyncDomain hipLaunchMemSyncDomain;
#define hipLaunchMemSyncDomainDefault cudaLaunchMemSyncDomainDefault
#define hipLaunchMemSyncDomainRemote cudaLaunchMemSyncDomainRemote
#endif
#if CUDA_VERSION >= CUDA_12000
typedef enum cudaGraphInstantiateResult hipGraphInstantiateResult;
#define hipGraphInstantiateSuccess cudaGraphInstantiateSuccess
@@ -1991,6 +2050,27 @@ inline static hipError_t hipMemPrefetchAsync(const void* dev_ptr, size_t count,
return hipCUDAErrorTohipError(cudaMemPrefetchAsync(dev_ptr, count, device, stream));
}
inline static hipError_t hipMemPrefetchAsync_v2(const void *dev_ptr, size_t count,
hipMemLocation location, unsigned int flags,
hipStream_t stream __dparm(0)) {
#if CUDA_VERSION >= 13000
return hipCUDAErrorTohipError(cudaMemPrefetchAsync(dev_ptr, count, location, flags, stream));
#else
return hipCUDAErrorTohipError(cudaMemPrefetchAsync_v2(dev_ptr, count, location, flags,
stream));
#endif
}
inline static hipError_t hipMemAdvise_v2(const void* dev_ptr, size_t count,
hipMemoryAdvise advice, hipMemLocation location) {
#if CUDA_VERSION >= 13000
return hipCUDAErrorTohipError(cudaMemAdvise(dev_ptr, count, advice , location));
#else
return hipCUDAErrorTohipError(cudaMemAdvise_v2(dev_ptr, count, advice , location));
#endif
}
inline static hipError_t hipMemRangeGetAttribute(void* data, size_t data_size,
hipMemRangeAttribute attribute,
const void* dev_ptr, size_t count) {
@@ -2091,6 +2171,10 @@ inline static hipError_t hipSetDevice(int device) {
return hipCUDAErrorTohipError(cudaSetDevice(device));
}
inline static hipError_t hipSetValidDevices(int* device_arr, int len) {
return hipCUDAErrorTohipError(cudaSetValidDevices(device_arr, len));
}
inline static hipError_t hipChooseDevice(int* device, const hipDeviceProp_t* prop) {
if (prop == NULL) {
@@ -2323,6 +2407,26 @@ inline static hipError_t hipMemcpy2DToArrayAsync(hipArray_t dst, size_t wOffset,
stream));
}
inline static 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) {
return hipCUDAErrorTohipError(cudaMemcpyBatchAsync(dsts, srcs, sizes, count, attrs, attrsIdxs,
numAttrs, failIdx, stream));
}
inline static hipError_t hipMemcpy3DBatchAsync(size_t numOps, hipMemcpy3DBatchOp *opList,
size_t *failIdx, unsigned long long flags,
hipStream_t stream) {
return hipCUDAErrorTohipError(cudaMemcpy3DBatchAsync(numOps, opList, failIdx, flags, stream));
}
inline static hipError_t hipMemcpy3DPeer(hipMemcpy3DPeerParms *p) {
return hipCUDAErrorTohipError(cudaMemcpy3DPeer(p));
}
inline static hipError_t hipMemcpy3DPeerAsync(hipMemcpy3DPeerParms *p, hipStream_t stream) {
return hipCUDAErrorTohipError(cudaMemcpy3DPeerAsync(p, stream));
}
__HIP_DEPRECATED inline static hipError_t hipMemcpyToArray(hipArray_t dst, size_t wOffset,
size_t hOffset, const void* src,
size_t count, hipMemcpyKind kind) {
@@ -2474,6 +2578,39 @@ inline static hipError_t hipMemset3DAsync(hipPitchedPtr pitchedDevPtr, int valu
return hipCUDAErrorTohipError(cudaMemset3DAsync(pitchedDevPtr, value, extent, stream));
}
inline static hipError_t hipMemsetD2D8(hipDeviceptr_t dst, size_t dstPitch, unsigned char value,
size_t width, size_t height) {
return hipCUResultTohipError(cuMemsetD2D8(dst, dstPitch, value, width, height));
}
inline static hipError_t hipMemsetD2D8Async(hipDeviceptr_t dst, size_t dstPitch,
unsigned char value, size_t width, size_t height,
hipStream_t stream) {
return hipCUResultTohipError(cuMemsetD2D8Async(dst, dstPitch, value, width, height, stream));
}
inline static hipError_t hipMemsetD2D16(hipDeviceptr_t dst, size_t dstPitch, unsigned short value,
size_t width, size_t height) {
return hipCUResultTohipError(cuMemsetD2D16(dst, dstPitch, value, width, height));
}
inline static hipError_t hipMemsetD2D16Async(hipDeviceptr_t dst, size_t dstPitch,
unsigned short value, size_t width, size_t height,
hipStream_t stream) {
return hipCUResultTohipError(cuMemsetD2D16Async(dst, dstPitch, value, width, height,
stream));
}
inline static hipError_t hipMemsetD2D32(hipDeviceptr_t dst, size_t dstPitch, unsigned int value,
size_t width, size_t height) {
return hipCUResultTohipError(cuMemsetD2D32(dst, dstPitch, value, width, height));
}
inline static hipError_t hipMemsetD2D32Async(hipDeviceptr_t dst, size_t dstPitch, unsigned int value,
size_t width, size_t height, hipStream_t stream) {
return hipCUResultTohipError(cuMemsetD2D32Async(dst, dstPitch, value, width, height,
stream));
}
inline static hipError_t hipGetDeviceProperties(hipDeviceProp_t* p_prop, int device) {
if (p_prop == NULL) {
return hipErrorInvalidValue;
@@ -2998,6 +3135,16 @@ inline static hipError_t hipMemUnmap(hipDeviceptr_t ptr, size_t size) {
}
#endif // CUDA_VERSION >= CUDA_10020
#if CUDA_VERSION >= CUDA_12000
inline static hipError_t hipMemGetHandleForAddressRange(void* handle, hipDeviceptr_t dptr,
size_t size,
hipMemRangeHandleType handleType,
unsigned long long flags) {
return hipCUResultTohipError(cuMemGetHandleForAddressRange(handle, dptr, size, handleType,
flags));
}
#endif
inline static hipError_t hipOccupancyMaxActiveBlocksPerMultiprocessor(int* numBlocks,
const void* func,
int blockSize,
@@ -3195,6 +3342,16 @@ inline static hipError_t hipStreamGetDevice(hipStream_t stream, hipDevice_t* dev
return hipCUResultTohipError(cuCtxPopCurrent(&context));
}
inline static hipError_t hipStreamSetAttribute(hipStream_t stream, hipStreamAttrID attr,
const hipStreamAttrValue* value) {
return hipCUDAErrorTohipError(cudaStreamSetAttribute(stream, attr, value));
}
inline static hipError_t hipStreamGetAttribute(hipStream_t stream, hipStreamAttrID attr,
hipStreamAttrValue* value_out) {
return hipCUDAErrorTohipError(cudaStreamGetAttribute(stream, attr, value_out));
}
inline static hipError_t hipDriverGetVersion(int* driverVersion) {
return hipCUDAErrorTohipError(cudaDriverGetVersion(driverVersion));
}
@@ -3426,6 +3583,10 @@ inline static hipError_t hipModuleLoad(hipModule_t* module, const char* fname) {
return hipCUResultTohipError(cuModuleLoad(module, fname));
}
inline static hipError_t hipModuleLoadFatBinary(hipModule_t* module, const void* fatbin) {
return hipCUResultTohipError(cuModuleLoadFatBinary(module, fatbin));
}
inline static hipError_t hipModuleUnload(hipModule_t hmod) {
return hipCUResultTohipError(cuModuleUnload(hmod));
}
@@ -3435,6 +3596,10 @@ inline static hipError_t hipModuleGetFunction(hipFunction_t* function, hipModule
return hipCUResultTohipError(cuModuleGetFunction(function, module, kname));
}
inline static hipError_t hipModuleGetFunctionCount(unsigned int* count, hipModule_t mod) {
return hipCUResultTohipError(cuModuleGetFunctionCount(count, mod));
}
inline static hipError_t hipModuleGetTexRef(hipTexRef* pTexRef, hipModule_t hmod, const char* name){
return hipCUResultTohipError(cuModuleGetTexRef(pTexRef, hmod, name));
}
@@ -3587,6 +3752,13 @@ inline static hipError_t hipGetChannelDesc(hipChannelFormatDesc* desc, hipArray_
return hipCUDAErrorTohipError(cudaGetChannelDesc(desc,array));
}
inline static hipError_t hipGetDriverEntryPoint(const char* symbol, void** funcPtr,
unsigned long long flags,
hipDriverEntryPointQueryResult* driverStatus)
{
return hipCUDAErrorTohipError(cudaGetDriverEntryPoint(symbol, funcPtr, flags, driverStatus));
}
inline static hipError_t hipLaunchCooperativeKernel(const void* f, dim3 gridDim, dim3 blockDim,
void** kernelParams, unsigned int sharedMemBytes,
hipStream_t stream) {
@@ -4708,12 +4880,19 @@ inline static void hipMemsetParamsToCUDAMemsetNodeParams(CUDA_MEMSET_NODE_PARAMS
}
inline static hipError_t hipDrvGraphAddMemsetNode(hipGraphNode_t* phGraphNode, hipGraph_t hGraph,
const hipGraphNode_t* dependencies, size_t numDependencies,
const hipMemsetParams* memsetParams, hipCtx_t ctx) {
CUDA_MEMSET_NODE_PARAMS cuMemsetParams;
hipMemsetParamsToCUDAMemsetNodeParams(&cuMemsetParams, memsetParams);
return hipCUResultTohipError(cuGraphAddMemsetNode(phGraphNode, hGraph, dependencies, numDependencies,
&cuMemsetParams, ctx));
const hipGraphNode_t* dependencies,
size_t numDependencies,
const hipMemsetParams* memsetParams, hipCtx_t ctx)
{
if (memsetParams == nullptr) {
return hipCUResultTohipError(
cuGraphAddMemsetNode(phGraphNode, hGraph, dependencies, numDependencies, nullptr, ctx));
} else {
CUDA_MEMSET_NODE_PARAMS cuMemsetParams;
hipMemsetParamsToCUDAMemsetNodeParams(&cuMemsetParams, memsetParams);
return hipCUResultTohipError(cuGraphAddMemsetNode(phGraphNode, hGraph, dependencies,
numDependencies, &cuMemsetParams, ctx));
}
}
inline static hipError_t hipDrvGraphAddMemcpyNode(hipGraphNode_t* phGraphNode, hipGraph_t hGraph,
@@ -4771,14 +4950,21 @@ inline static hipError_t hipDrvGraphExecMemcpyNodeSetParams(hipGraphExec_t hGrap
}
}
inline static hipError_t hipDrvGraphExecMemsetNodeSetParams(
hipGraphExec_t hGraphExec, hipGraphNode_t hNode, const hipMemsetParams* memsetParams,
hipCtx_t ctx) {
CUDA_MEMSET_NODE_PARAMS cuMemsetParams;
hipMemsetParamsToCUDAMemsetNodeParams(&cuMemsetParams, memsetParams);
return hipCUResultTohipError(
cuGraphExecMemsetNodeSetParams(hGraphExec, hNode, &cuMemsetParams, ctx));
}
inline static hipError_t hipDrvGraphExecMemsetNodeSetParams(hipGraphExec_t hGraphExec,
hipGraphNode_t hNode,
const hipMemsetParams* memsetParams,
hipCtx_t ctx)
{
if (memsetParams == nullptr) {
return hipCUResultTohipError(
cuGraphExecMemsetNodeSetParams(hGraphExec, hNode, nullptr, ctx));
} else {
CUDA_MEMSET_NODE_PARAMS cuMemsetParams;
hipMemsetParamsToCUDAMemsetNodeParams(&cuMemsetParams, memsetParams);
return hipCUResultTohipError(
cuGraphExecMemsetNodeSetParams(hGraphExec, hNode, &cuMemsetParams, ctx));
}
}
#endif
#if CUDA_VERSION >= CUDA_11040
@@ -4830,6 +5016,10 @@ inline static hipError_t hipMemcpy2DArrayToArray(hipArray_t dst, size_t wOffsetD
inline static hipError_t hipSetValidDevices(int* device_arr, int len) {
return hipCUDAErrorTohipError(cudaSetValidDevices(device_arr, len));
}
#endif //__CUDACC__
#endif // HIP_INCLUDE_HIP_NVIDIA_DETAIL_HIP_RUNTIME_API_H