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

SWDEV-332969 - [ABI Break]Substitute hipArray* with hipArray_t

Browse Source 
- hipArray will be an internal struct from rocm6.0

Signed-off-by: sdashmiz <shadi.dashmiz@amd.com>
Change-Id: Icf97fe96b87be8532098cd7f9ceaad099f99c9b9


[ROCm/hipother commit: 452c1060f3]
This commit is contained in:
sdashmiz
2023-07-25 14:29:18 -04:00
committed by Rakesh Roy
parent 63029a9739
commit dd373854f3
1 changed files with 257 additions and 35 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
+257 -35
View File
@@ -254,7 +254,6 @@ inline static CUresourcetype hipResourcetype_enumToCUresourcetype(
#define hipStreamPerThread ((cudaStream_t)2)
#define hipTexRef CUtexref
#define hiparray CUarray
typedef CUmipmappedArray hipmipmappedArray;
typedef cudaMipmappedArray_t hipMipmappedArray_t;
@@ -438,7 +437,6 @@ typedef enum cudaDeviceP2PAttr hipDeviceP2PAttr;
typedef CUmodule hipModule_t;
typedef CUfunction hipFunction_t;
typedef CUdeviceptr hipDeviceptr_t;
typedef struct cudaArray hipArray;
typedef struct cudaArray* hipArray_t;
typedef struct cudaArray* hipArray_const_t;
typedef struct cudaFuncAttributes hipFuncAttributes;
@@ -446,8 +444,209 @@ typedef struct cudaLaunchParams hipLaunchParams;
typedef CUDA_LAUNCH_PARAMS hipFunctionLaunchParams;
#define hipFunction_attribute CUfunction_attribute
#define hipPointer_attribute CUpointer_attribute
#define hip_Memcpy2D CUDA_MEMCPY2D
#define HIP_MEMCPY3D CUDA_MEMCPY3D
typedef struct HIP_RESOURCE_DESC_st
{
hipResourcetype resType; /**< Resource type */
union {
struct {
hipArray_t hArray; /**< HIP array */
} array;
struct {
hipMipmappedArray_t hMipmappedArray; /**< HIP mipmapped array */
} mipmap;
struct {
hipDeviceptr_t devPtr; /**< Device pointer */
hipArray_Format format; /**< Array format */
unsigned int numChannels; /**< Channels per array element */
size_t sizeInBytes; /**< Size in bytes */
} linear;
struct {
hipDeviceptr_t devPtr; /**< Device pointer */
hipArray_Format format; /**< Array format */
unsigned int numChannels; /**< Channels per array element */
size_t width; /**< Width of the array in elements */
size_t height; /**< Height of the array in elements */
size_t pitchInBytes; /**< Pitch between two rows in bytes */
} pitch2D;
struct {
int reserved[32];
} reserved;
} res;
unsigned int flags; /**< Flags (must be zero) */
} HIP_RESOURCE_DESC;
static inline CUDA_RESOURCE_DESC* hipResourceDesTocudaResourceDes(const HIP_RESOURCE_DESC* p){
CUDA_RESOURCE_DESC a;
switch (p->resType) {
case HIP_RESOURCE_TYPE_ARRAY:
a.resType = CU_RESOURCE_TYPE_ARRAY;
case HIP_RESOURCE_TYPE_MIPMAPPED_ARRAY:
a.resType = CU_RESOURCE_TYPE_MIPMAPPED_ARRAY;
case HIP_RESOURCE_TYPE_LINEAR:
a.resType = CU_RESOURCE_TYPE_LINEAR;
case HIP_RESOURCE_TYPE_PITCH2D:
a.resType = CU_RESOURCE_TYPE_PITCH2D;
default:
a.resType = CU_RESOURCE_TYPE_ARRAY;
}
a.res.array.hArray = (CUarray)p->res.array.hArray;
a.res.mipmap.hMipmappedArray = (CUmipmappedArray)p->res.mipmap.hMipmappedArray;
a.res.linear.devPtr = p->res.linear.devPtr;
a.res.linear.format = p->res.linear.format;
a.res.linear.numChannels = p->res.linear.numChannels;
a.res.linear.sizeInBytes = p->res.linear.sizeInBytes;
a.res.pitch2D.devPtr = p->res.pitch2D.devPtr;
a.res.pitch2D.numChannels = p->res.pitch2D.numChannels;
a.res.pitch2D.format = p->res.pitch2D.format;
a.res.pitch2D.width = p->res.pitch2D.width;
a.res.pitch2D.height = p->res.pitch2D.height;
a.res.pitch2D.pitchInBytes = p->res.pitch2D.pitchInBytes;
a.flags = p->flags;
return &a;
}
typedef struct hip_Memcpy2D {
size_t srcXInBytes;
size_t srcY;
hipMemoryType srcMemoryType;
const void* srcHost;
hipDeviceptr_t srcDevice;
hipArray_t srcArray;
size_t srcPitch;
size_t dstXInBytes;
size_t dstY;
hipMemoryType dstMemoryType;
void* dstHost;
hipDeviceptr_t dstDevice;
hipArray_t dstArray;
size_t dstPitch;
size_t WidthInBytes;
size_t Height;
} hip_Memcpy2D;
typedef struct HIP_MEMCPY3D {
unsigned int srcXInBytes;
unsigned int srcY;
unsigned int srcZ;
unsigned int srcLOD;
hipMemoryType srcMemoryType;
const void* srcHost;
hipDeviceptr_t srcDevice;
hipArray_t srcArray;
unsigned int srcPitch;
unsigned int srcHeight;
unsigned int dstXInBytes;
unsigned int dstY;
unsigned int dstZ;
unsigned int dstLOD;
hipMemoryType dstMemoryType;
void* dstHost;
hipDeviceptr_t dstDevice;
hipArray_t dstArray;
unsigned int dstPitch;
unsigned int dstHeight;
unsigned int WidthInBytes;
unsigned int Height;
unsigned int Depth;
} HIP_MEMCPY3D;
static inline void hipMemcpy3DTocudaMemcpy3D(CUDA_MEMCPY3D &a, const HIP_MEMCPY3D* p){
a.srcXInBytes = (size_t)p->srcXInBytes;
a.srcY = (size_t)p->srcY;
a.srcZ = (size_t)p->srcZ;
a.srcLOD = (size_t)p->srcLOD;
switch (p->srcMemoryType) {
case hipMemoryTypeHost:
a.srcMemoryType = CU_MEMORYTYPE_HOST;
break;
case hipMemoryTypeDevice:
a.srcMemoryType = CU_MEMORYTYPE_DEVICE;
break;
case hipMemoryTypeArray:
a.srcMemoryType = CU_MEMORYTYPE_ARRAY;
break;
default:
a.srcMemoryType = CU_MEMORYTYPE_UNIFIED;
}
a.srcHost = p->srcHost;
a.srcDevice =(CUdeviceptr)p->srcDevice;
a.srcArray = (CUarray)p->srcArray;
a.reserved0 = nullptr;
a.srcPitch = (size_t)p->srcPitch;
a.srcHeight = (size_t)p->srcHeight;
a.dstXInBytes = (size_t)p->dstXInBytes;
a.dstY = (size_t)p->dstY;
a.dstZ = (size_t)p->dstZ;
a.dstLOD = (size_t)p->dstLOD;
switch (p->dstMemoryType) {
case hipMemoryTypeHost:
a.dstMemoryType = CU_MEMORYTYPE_HOST;
break;
case hipMemoryTypeDevice:
a.dstMemoryType = CU_MEMORYTYPE_DEVICE;
break;
case hipMemoryTypeArray:
a.dstMemoryType = CU_MEMORYTYPE_ARRAY;
break;
default:
a.dstMemoryType = CU_MEMORYTYPE_UNIFIED;
}
a.dstHost = p->dstHost;
a.dstDevice = (CUdeviceptr)p->dstDevice;
a.dstArray = (CUarray)p->dstArray;
a.reserved1 = nullptr;
a.dstPitch = (size_t)p->dstPitch;
a.dstHeight = (size_t)p->dstHeight;
a.WidthInBytes = (size_t)p->WidthInBytes;
a.Height = (size_t)p->Height;
a.Depth = (size_t)p->Depth;
}
static inline void hipMemcpy2DTocudaMemcpy2D(CUDA_MEMCPY2D &a, const hip_Memcpy2D* p){
a.srcXInBytes = (size_t)p->srcXInBytes;
a.srcY = (size_t)p->srcY;
switch (p->srcMemoryType) {
case hipMemoryTypeHost:
a.srcMemoryType = CU_MEMORYTYPE_HOST;
break;
case hipMemoryTypeDevice:
a.srcMemoryType = CU_MEMORYTYPE_DEVICE;
break;
case hipMemoryTypeArray:
a.srcMemoryType = CU_MEMORYTYPE_ARRAY;
break;
default:
a.srcMemoryType = CU_MEMORYTYPE_UNIFIED;
}
a.srcHost = p->srcHost;
a.srcDevice = (CUdeviceptr)p->srcDevice;
a.srcArray = (CUarray)p->srcArray;
a.srcPitch = (size_t)p->srcPitch;
a.dstXInBytes = (size_t)p->dstXInBytes;
a.dstY = (size_t)p->dstY;
switch (p->dstMemoryType) {
case hipMemoryTypeHost:
a.dstMemoryType = CU_MEMORYTYPE_HOST;
break;
case hipMemoryTypeDevice:
a.dstMemoryType = CU_MEMORYTYPE_DEVICE;
break;
case hipMemoryTypeArray:
a.dstMemoryType = CU_MEMORYTYPE_ARRAY;
break;
default:
a.dstMemoryType = CU_MEMORYTYPE_UNIFIED;
}
a.dstHost = p->dstHost;
a.dstDevice = (CUdeviceptr)p->dstDevice;
a.dstArray = (CUarray)p->dstArray;
a.dstPitch = (size_t)p->dstPitch;
a.WidthInBytes = (size_t)p->WidthInBytes;
a.Height = (size_t)p->Height;
}
#define hipMemcpy3DParms cudaMemcpy3DParms
#define hipArrayDefault cudaArrayDefault
#define hipArrayLayered cudaArrayLayered
@@ -507,7 +706,6 @@ typedef struct cudaChannelFormatDesc hipChannelFormatDesc;
typedef struct cudaResourceDesc hipResourceDesc;
typedef struct cudaTextureDesc hipTextureDesc;
typedef struct cudaResourceViewDesc hipResourceViewDesc;
typedef CUDA_RESOURCE_DESC HIP_RESOURCE_DESC;
typedef CUDA_TEXTURE_DESC HIP_TEXTURE_DESC;
typedef CUDA_RESOURCE_VIEW_DESC HIP_RESOURCE_VIEW_DESC;
// adding code for hipmemSharedConfig
@@ -1568,18 +1766,18 @@ inline static hipError_t hipMallocManaged(void** ptr, size_t size, unsigned int
return hipCUDAErrorTohipError(cudaMallocManaged(ptr, size, flags));
}
inline static hipError_t hipMallocArray(hipArray** array, const hipChannelFormatDesc* desc,
inline static hipError_t hipMallocArray(hipArray_t* array, const hipChannelFormatDesc* desc,
size_t width, size_t height __dparm(0),
unsigned int flags __dparm(hipArrayDefault)) {
return hipCUDAErrorTohipError(cudaMallocArray(array, desc, width, height, flags));
}
inline static hipError_t hipMalloc3DArray(hipArray** array, const hipChannelFormatDesc* desc,
inline static hipError_t hipMalloc3DArray(hipArray_t* array, const hipChannelFormatDesc* desc,
hipExtent extent, unsigned int flags) {
return hipCUDAErrorTohipError(cudaMalloc3DArray(array, desc, extent, flags));
}
inline static hipError_t hipFreeArray(hipArray* array) {
inline static hipError_t hipFreeArray(hipArray_t array) {
return hipCUDAErrorTohipError(cudaFreeArray(array));
}
@@ -1593,7 +1791,7 @@ inline static hipError_t hipMipmappedArrayDestroy(hipmipmappedArray hMipmappedAr
return hipCUResultTohipError(cuMipmappedArrayDestroy(hMipmappedArray));
}
inline static hipError_t hipMipmappedArrayGetLevel(hiparray* pLevelArray,
inline static hipError_t hipMipmappedArrayGetLevel(hipArray_t* pLevelArray,
hipmipmappedArray hMipMappedArray,
unsigned int level) {
return hipCUResultTohipError(cuMipmappedArrayGetLevel((CUarray*)pLevelArray, hMipMappedArray, level));
@@ -1763,11 +1961,23 @@ inline static hipError_t hipMemcpy2D(void* dst, size_t dpitch, const void* src,
}
inline static hipError_t hipMemcpyParam2D(const hip_Memcpy2D* pCopy) {
return hipCUResultTohipError(cuMemcpy2D(pCopy));
if(pCopy == nullptr) {
return hipCUResultTohipError(cuMemcpy2D(nullptr));
} else {
CUDA_MEMCPY2D cudaCopy = {0};
hipMemcpy2DTocudaMemcpy2D(cudaCopy, pCopy);
return hipCUResultTohipError(cuMemcpy2D((const CUDA_MEMCPY2D*)&cudaCopy));
}
}
inline static hipError_t hipMemcpyParam2DAsync(const hip_Memcpy2D* pCopy, hipStream_t stream __dparm(0)) {
return hipCUResultTohipError(cuMemcpy2DAsync(pCopy, stream));
if(pCopy == nullptr) {
return hipCUResultTohipError(cuMemcpy2DAsync(nullptr, stream));
} else {
CUDA_MEMCPY2D cudaCopy = {0};
hipMemcpy2DTocudaMemcpy2D(cudaCopy, pCopy);
return hipCUResultTohipError(cuMemcpy2DAsync((const CUDA_MEMCPY2D*)&cudaCopy, stream));
}
}
inline static hipError_t hipMemcpy3D(const struct hipMemcpy3DParms *p) {
@@ -1778,12 +1988,24 @@ inline static hipError_t hipMemcpy3DAsync(const struct hipMemcpy3DParms *p, hipS
return hipCUDAErrorTohipError(cudaMemcpy3DAsync(p, stream));
}
inline static hipError_t hipDrvMemcpy3D(const HIP_MEMCPY3D* pCopy) {
return hipCUResultTohipError(cuMemcpy3D(pCopy));
inline static hipError_t hipDrvMemcpy3D(const HIP_MEMCPY3D* pcopy) {
if(pcopy == nullptr) {
return hipCUResultTohipError(cuMemcpy3D(nullptr));
} else {
CUDA_MEMCPY3D cudaCopy = {0};
hipMemcpy3DTocudaMemcpy3D(cudaCopy, pcopy);
return hipCUResultTohipError(cuMemcpy3D((const CUDA_MEMCPY3D*)&cudaCopy));
}
}
inline static hipError_t hipDrvMemcpy3DAsync(const HIP_MEMCPY3D* pCopy, hipStream_t stream) {
return hipCUResultTohipError(cuMemcpy3DAsync(pCopy, stream));
inline static hipError_t hipDrvMemcpy3DAsync(const HIP_MEMCPY3D *pcopy, hipStream_t stream) {
if(pcopy == nullptr) {
return hipCUResultTohipError(cuMemcpy3DAsync(nullptr, stream));
} else {
CUDA_MEMCPY3D cudaCopy = {0};
hipMemcpy3DTocudaMemcpy3D(cudaCopy, pcopy);
return hipCUResultTohipError(cuMemcpy3DAsync((const CUDA_MEMCPY3D*)&cudaCopy, stream));
}
}
inline static hipError_t hipMemcpy2DAsync(void* dst, size_t dpitch, const void* src, size_t spitch,
@@ -1793,7 +2015,7 @@ inline static hipError_t hipMemcpy2DAsync(void* dst, size_t dpitch, const void*
kind, stream));
}
inline static hipError_t hipMemcpy2DFromArray(void* dst, size_t dpitch, hipArray* src,
inline static hipError_t hipMemcpy2DFromArray(void* dst, size_t dpitch, hipArray_t src,
size_t wOffset, size_t hOffset, size_t width,
size_t height, hipMemcpyKind kind) {
return hipCUDAErrorTohipError(cudaMemcpy2DFromArray(dst, dpitch, src, wOffset, hOffset, width,
@@ -1801,7 +2023,7 @@ inline static hipError_t hipMemcpy2DFromArray(void* dst, size_t dpitch, hipArray
kind));
}
inline static hipError_t hipMemcpy2DFromArrayAsync(void* dst, size_t dpitch, hipArray* src,
inline static hipError_t hipMemcpy2DFromArrayAsync(void* dst, size_t dpitch, hipArray_t src,
size_t wOffset, size_t hOffset, size_t width,
size_t height, hipMemcpyKind kind,
hipStream_t stream) {
@@ -1811,14 +2033,14 @@ inline static hipError_t hipMemcpy2DFromArrayAsync(void* dst, size_t dpitch, hip
stream));
}
inline static hipError_t hipMemcpy2DToArray(hipArray* dst, size_t wOffset, size_t hOffset,
inline static hipError_t hipMemcpy2DToArray(hipArray_t dst, size_t wOffset, size_t hOffset,
const void* src, size_t spitch, size_t width,
size_t height, hipMemcpyKind kind) {
return hipCUDAErrorTohipError(cudaMemcpy2DToArray(dst, wOffset, hOffset, src, spitch, width,
height, kind));
}
inline static hipError_t hipMemcpy2DToArrayAsync(hipArray* dst, size_t wOffset, size_t hOffset,
inline static hipError_t hipMemcpy2DToArrayAsync(hipArray_t dst, size_t wOffset, size_t hOffset,
const void* src, size_t spitch, size_t width,
size_t height, hipMemcpyKind kind,
hipStream_t stream) {
@@ -1828,7 +2050,7 @@ inline static hipError_t hipMemcpy2DToArrayAsync(hipArray* dst, size_t wOffset,
stream));
}
__HIP_DEPRECATED inline static hipError_t hipMemcpyToArray(hipArray* dst, size_t wOffset,
__HIP_DEPRECATED inline static hipError_t hipMemcpyToArray(hipArray_t dst, size_t wOffset,
size_t hOffset, const void* src,
size_t count, hipMemcpyKind kind) {
return hipCUDAErrorTohipError(
@@ -1842,12 +2064,12 @@ __HIP_DEPRECATED inline static hipError_t hipMemcpyFromArray(void* dst, hipArray
kind));
}
inline static hipError_t hipMemcpyAtoH(void* dst, hipArray* srcArray, size_t srcOffset,
inline static hipError_t hipMemcpyAtoH(void* dst, hipArray_t srcArray, size_t srcOffset,
size_t count) {
return hipCUResultTohipError(cuMemcpyAtoH(dst, (CUarray)srcArray, srcOffset, count));
}
inline static hipError_t hipMemcpyHtoA(hipArray* dstArray, size_t dstOffset, const void* srcHost,
inline static hipError_t hipMemcpyHtoA(hipArray_t dstArray, size_t dstOffset, const void* srcHost,
size_t count) {
return hipCUResultTohipError(cuMemcpyHtoA((CUarray)dstArray, dstOffset, srcHost, count));
}
@@ -3241,7 +3463,7 @@ inline static hipError_t hipTexObjectCreate(hipTextureObject_t* pTexObject,
const HIP_RESOURCE_DESC* pResDesc,
const HIP_TEXTURE_DESC* pTexDesc,
const HIP_RESOURCE_VIEW_DESC* pResViewDesc) {
return hipCUResultTohipError(cuTexObjectCreate((CUtexObject*)pTexObject, pResDesc, pTexDesc, pResViewDesc));
return hipCUResultTohipError(cuTexObjectCreate((CUtexObject*)pTexObject,(CUDA_RESOURCE_DESC*)pResDesc, pTexDesc, pResViewDesc));
}
inline static hipError_t hipTexObjectDestroy(hipTextureObject_t texObject) {
@@ -3249,7 +3471,7 @@ inline static hipError_t hipTexObjectDestroy(hipTextureObject_t texObject) {
}
inline static hipError_t hipTexObjectGetResourceDesc(HIP_RESOURCE_DESC* pResDesc, hipTextureObject_t texObject) {
return hipCUResultTohipError(cuTexObjectGetResourceDesc(pResDesc, (CUtexObject)texObject));
return hipCUResultTohipError(cuTexObjectGetResourceDesc((CUDA_RESOURCE_DESC*)pResDesc, (CUtexObject)texObject));
}
inline static hipError_t hipTexObjectGetResourceViewDesc(HIP_RESOURCE_VIEW_DESC* pResViewDesc, hipTextureObject_t texObject) {
@@ -3284,35 +3506,35 @@ __HIP_DEPRECATED inline static hipError_t hipTexRefSetFlags(hipTexRef hTexRef, u
return hipCUResultTohipError(cuTexRefSetFlags(hTexRef,Flags));
}
__HIP_DEPRECATED inline static hipError_t hipTexRefSetArray(hipTexRef hTexRef, hiparray hArray, unsigned int Flags){
return hipCUResultTohipError(cuTexRefSetArray(hTexRef,hArray,Flags));
__HIP_DEPRECATED inline static hipError_t hipTexRefSetArray(hipTexRef hTexRef, hipArray_t hArray, unsigned int Flags){
return hipCUResultTohipError(cuTexRefSetArray(hTexRef,(CUarray)hArray,Flags));
}
inline static hipError_t hipArrayCreate(hiparray* pHandle, const HIP_ARRAY_DESCRIPTOR* pAllocateArray){
return hipCUResultTohipError(cuArrayCreate(pHandle, pAllocateArray));
inline static hipError_t hipArrayCreate(hipArray_t* pHandle, const HIP_ARRAY_DESCRIPTOR* pAllocateArray){
return hipCUResultTohipError(cuArrayCreate((CUarray*)pHandle, pAllocateArray));
}
inline static hipError_t hipArrayDestroy(hiparray hArray){
return hipCUResultTohipError(cuArrayDestroy(hArray));
inline static hipError_t hipArrayDestroy(hipArray_t hArray){
return hipCUResultTohipError(cuArrayDestroy((CUarray)hArray));
}
inline static hipError_t hipArray3DCreate(hiparray* pHandle,
inline static hipError_t hipArray3DCreate(hipArray_t* pHandle,
const HIP_ARRAY3D_DESCRIPTOR* pAllocateArray){
return hipCUResultTohipError(cuArray3DCreate(pHandle, pAllocateArray));
return hipCUResultTohipError(cuArray3DCreate((CUarray*)pHandle, pAllocateArray));
}
inline static hipError_t hipArrayGetInfo(hipChannelFormatDesc* desc, hipExtent* extent,
unsigned int* flags, hipArray* array) {
unsigned int* flags, hipArray_t array) {
return hipCUDAErrorTohipError(cudaArrayGetInfo(desc, extent, flags, array));
}
inline static hipError_t hipArrayGetDescriptor(HIP_ARRAY_DESCRIPTOR* pArrayDescriptor,
hipArray* array) {
hipArray_t array) {
return hipCUResultTohipError(cuArrayGetDescriptor(pArrayDescriptor, (CUarray)array));
}
inline static hipError_t hipArray3DGetDescriptor(HIP_ARRAY3D_DESCRIPTOR* pArrayDescriptor,
hipArray* array) {
hipArray_t array) {
return hipCUResultTohipError(cuArray3DGetDescriptor(pArrayDescriptor, (CUarray)array));
}