diff --git a/hipamd/include/hip/hcc_detail/channel_descriptor.h b/hipamd/include/hip/hcc_detail/channel_descriptor.h index 38acff9951..a69558c8e4 100644 --- a/hipamd/include/hip/hcc_detail/channel_descriptor.h +++ b/hipamd/include/hip/hcc_detail/channel_descriptor.h @@ -29,8 +29,14 @@ THE SOFTWARE. #ifdef __cplusplus +#if __HIP_VDI__ +extern "C" { +#endif HIP_PUBLIC_API hipChannelFormatDesc hipCreateChannelDesc(int x, int y, int z, int w, hipChannelFormatKind f); +#if __HIP_VDI__ +} +#endif static inline hipChannelFormatDesc hipCreateChannelDescHalf() { int e = (int)sizeof(unsigned short) * 8; diff --git a/hipamd/include/hip/hcc_detail/hip_runtime_api.h b/hipamd/include/hip/hcc_detail/hip_runtime_api.h index 2a781154e1..d6b66090cd 100644 --- a/hipamd/include/hip/hcc_detail/hip_runtime_api.h +++ b/hipamd/include/hip/hcc_detail/hip_runtime_api.h @@ -3314,6 +3314,172 @@ hipError_t hipLaunchKernel(const void* function_address, size_t sharedMemBytes __dparm(0), hipStream_t stream __dparm(0)); +#if __HIP_VDI__ +hipError_t hipBindTexture( + size_t* offset, + const textureReference* tex, + const void* devPtr, + const hipChannelFormatDesc* desc, + size_t size = UINT_MAX); + +hipError_t hipBindTexture2D( + size_t* offset, + const textureReference* tex, + const void* devPtr, + const hipChannelFormatDesc* desc, + size_t width, + size_t height, + size_t pitch); + +hipError_t hipBindTextureToArray( + const textureReference* tex, + hipArray_const_t array, + const hipChannelFormatDesc* desc); + +hipError_t hipBindTextureToMipmappedArray( + const textureReference* tex, + hipMipmappedArray_const_t mipmappedArray, + const hipChannelFormatDesc* desc); + +hipError_t hipGetTextureAlignmentOffset( + size_t* offset, + const textureReference* texref); + +hipError_t hipGetTextureReference( + const textureReference** texref, + const void* symbol); + +hipError_t hipUnbindTexture(const textureReference* tex); + +hipError_t hipCreateTextureObject( + hipTextureObject_t* pTexObject, + const hipResourceDesc* pResDesc, + const hipTextureDesc* pTexDesc, + const hipResourceViewDesc* pResViewDesc); + +hipError_t hipDestroyTextureObject(hipTextureObject_t textureObject); + +hipError_t hipGetChannelDesc( + hipChannelFormatDesc* desc, + hipArray_const_t array); + +hipError_t hipGetTextureObjectResourceDesc( + hipResourceDesc* pResDesc, + hipTextureObject_t textureObject); + +hipError_t hipGetTextureObjectResourceViewDesc( + hipResourceViewDesc* pResViewDesc, + hipTextureObject_t textureObject); + +hipError_t hipGetTextureObjectTextureDesc( + hipTextureDesc* pTexDesc, + hipTextureObject_t textureObject); + +hipError_t hipTexRefGetAddress( + hipDeviceptr_t* dev_ptr, + const textureReference* texRef); + +hipError_t hipTexRefGetAddressMode( + hipTextureAddressMode* pam, + const textureReference* texRef, + int dim); + +hipError_t hipTexRefGetFilterMode( + hipTextureFilterMode* pfm, + const textureReference* texRef); + +hipError_t hipTexRefGetFlags( + unsigned int* pFlags, + const textureReference* texRef); + +hipError_t hipTexRefGetFormat( + hipArray_Format* pFormat, + int* pNumChannels, + const textureReference* texRef); + +hipError_t hipTexRefGetMaxAnisotropy( + int* pmaxAnsio, + const textureReference* texRef); + +hipError_t hipTexRefGetMipmapFilterMode( + hipTextureFilterMode* pfm, + const textureReference* texRef); + +hipError_t hipTexRefGetMipmapLevelBias( + float* pbias, + const textureReference* texRef); + +hipError_t hipTexRefGetMipmapLevelClamp( + float* pminMipmapLevelClamp, + float* pmaxMipmapLevelClamp, + const textureReference* texRef); + +hipError_t hipTexRefGetMipMappedArray( + hipMipmappedArray_t* pArray, + const textureReference* texRef); + +hipError_t hipTexRefSetAddress( + size_t* ByteOffset, + textureReference* texRef, + hipDeviceptr_t dptr, + size_t bytes); + +hipError_t hipTexRefSetAddress2D( + textureReference* texRef, + const HIP_ARRAY_DESCRIPTOR* desc, + hipDeviceptr_t dptr, + size_t Pitch); + +hipError_t hipTexRefSetAddressMode( + textureReference* texRef, + int dim, + hipTextureAddressMode am); + +hipError_t hipTexRefSetArray( + textureReference* tex, + hipArray_const_t array, + unsigned int flags); + +hipError_t hipTexRefSetBorderColor( + textureReference* texRef, + float* pBorderColor); + +hipError_t hipTexRefSetFilterMode( + textureReference* texRef, + hipTextureFilterMode fm); + +hipError_t hipTexRefSetFlags( + textureReference* texRef, + unsigned int Flags); + +hipError_t hipTexRefSetFormat( + textureReference* texRef, + hipArray_Format fmt, + int NumPackedComponents); + +hipError_t hipTexRefSetMaxAnisotropy( + textureReference* texRef, + unsigned int maxAniso); + +hipError_t hipTexRefSetMipmapFilterMode( + textureReference* texRef, + hipTextureFilterMode fm); + +hipError_t hipTexRefSetMipmapLevelBias( + textureReference* texRef, + float bias); + +hipError_t hipTexRefSetMipmapLevelClamp( + textureReference* texRef, + float minMipMapLevelClamp, + float maxMipMapLevelClamp); + +hipError_t hipTexRefSetMipmappedArray( + textureReference* texRef, + hipMipmappedArray* mipmappedArray, + unsigned int Flags); +#endif + /** * @} */ @@ -3362,14 +3528,7 @@ const char* hipKernelNameRef(const hipFunction_t f); class TlsData; -#if __HIP_VDI__ -hipError_t hipBindTexture( - size_t* offset, - const textureReference* tex, - const void* devPtr, - const hipChannelFormatDesc* desc, - size_t size = UINT_MAX); -#else +#if !__HIP_VDI__ hipError_t hipBindTexture(size_t* offset, textureReference* tex, const void* devPtr, const hipChannelFormatDesc* desc, size_t size = UINT_MAX); #endif @@ -3426,16 +3585,7 @@ hipError_t hipBindTexture(size_t* offset, struct texture& tex, #endif // C API -#if __HIP_VDI__ -hipError_t hipBindTexture2D( - size_t* offset, - const textureReference* tex, - const void* devPtr, - const hipChannelFormatDesc* desc, - size_t width, - size_t height, - size_t pitch); -#else +#if !__HIP_VDI__ hipError_t hipBindTexture2D(size_t* offset, textureReference* tex, const void* devPtr, const hipChannelFormatDesc* desc, size_t width, size_t height, size_t pitch); @@ -3466,12 +3616,7 @@ hipError_t hipBindTexture2D(size_t* offset, struct texture& te #endif // C API -#if __HIP_VDI__ -hipError_t hipBindTextureToArray( - const textureReference* tex, - hipArray_const_t array, - const hipChannelFormatDesc* desc); -#else +#if !__HIP_VDI__ hipError_t hipBindTextureToArray(textureReference* tex, hipArray_const_t array, const hipChannelFormatDesc* desc); #endif @@ -3508,9 +3653,11 @@ inline static hipError_t hipBindTextureToArray(struct texture #endif // C API +#if !__HIP_VDI__ hipError_t hipBindTextureToMipmappedArray(const textureReference* tex, hipMipmappedArray_const_t mipmappedArray, const hipChannelFormatDesc* desc); +#endif #if !__HIP_VDI__ template @@ -3566,7 +3713,9 @@ inline hipError_t hipExtLaunchMultiKernelMultiDevice(hipLaunchParams* launchPara * * @return #hipSuccess **/ +#if !__HIP_VDI__ hipError_t hipUnbindTexture(const textureReference* tex); +#endif #if !__HIP_VDI__ extern hipError_t ihipUnbindTextureImpl(const hipTextureObject_t& textureObject); @@ -3579,6 +3728,7 @@ hipError_t hipUnbindTexture(struct texture& tex) { } #endif +#if !__HIP_VDI__ hipError_t hipGetChannelDesc(hipChannelFormatDesc* desc, hipArray_const_t array); hipError_t hipGetTextureAlignmentOffset(size_t* offset, const textureReference* texref); hipError_t hipGetTextureReference(const textureReference** texref, const void* symbol); @@ -3616,6 +3766,7 @@ hipError_t hipTexRefGetAddress(hipDeviceptr_t* dev_ptr, textureReference tex); hipError_t hipTexRefSetAddress2D(textureReference* tex, const HIP_ARRAY_DESCRIPTOR* desc, hipDeviceptr_t devPtr, size_t pitch); +#endif hipError_t hipCreateSurfaceObject(hipSurfaceObject_t* pSurfObject, const hipResourceDesc* pResDesc); diff --git a/hipamd/vdi/hip_hcc.def.in b/hipamd/vdi/hip_hcc.def.in index 57d3a80719..c6e4a00200 100644 --- a/hipamd/vdi/hip_hcc.def.in +++ b/hipamd/vdi/hip_hcc.def.in @@ -173,31 +173,45 @@ hipRegisterActivityCallback hipRemoveActivityCallback hipApiName hipKernelNameRef -hipCreateTextureObject -hipDestroyTextureObject -hipGetTextureObjectResourceDesc -hipGetTextureObjectResourceViewDesc -hipGetTextureObjectTextureDesc hipBindTexture hipBindTexture2D hipBindTextureToArray hipBindTextureToMipmappedArray -hipUnbindTexture -hipGetChannelDesc hipGetTextureAlignmentOffset hipGetTextureReference -hipTexRefSetFormat -hipTexRefSetFlags -hipTexRefSetFilterMode -hipTexRefSetAddressMode -hipTexRefSetArray -hipTexRefSetAddress -hipTexRefSetAddress2D +hipUnbindTexture +hipCreateChannelDesc +hipCreateTextureObject +hipDestroyTextureObject +hipGetChannelDesc +hipGetTextureObjectResourceDesc +hipGetTextureObjectResourceViewDesc +hipGetTextureObjectTextureDesc hipTexRefGetAddress hipTexRefGetAddressMode hipTexRefGetArray +hipTexRefGetBorderColor +hipTexRefGetFilterMode +hipTexRefGetFlags +hipTexRefGetFormat +hipTexRefGetMaxAnisotropy +hipTexRefGetMipmapFilterMode +hipTexRefGetMipmapLevelBias +hipTexRefGetMipmapLevelClamp +hipTexRefGetMipmappedArray +hipTexRefSetAddress +hipTexRefSetAddress2D +hipTexRefSetAddressMode hipTexRefSetArray -hipCreateChannelDesc +hipTexRefSetBorderColor +hipTexRefSetFilterMode +hipTexRefSetFlags +hipTexRefSetFormat +hipTexRefSetMaxAnisotropy +hipTexRefSetMipmapFilterMode +hipTexRefSetMipmapLevelBias +hipTexRefSetMipmapLevelClamp +hipTexRefSetMipmappedArray hipProfilerStart hipProfilerStop hipHccGetAccelerator diff --git a/hipamd/vdi/hip_hcc.map.in b/hipamd/vdi/hip_hcc.map.in index 70109c4943..c26e79fc70 100644 --- a/hipamd/vdi/hip_hcc.map.in +++ b/hipamd/vdi/hip_hcc.map.in @@ -186,34 +186,48 @@ global: hiprtcGetErrorString; hiprtcAddNameExpression; hiprtcVersion; + hipBindTexture; + hipBindTexture2D; + hipBindTextureToArray; + hipBindTextureToMipmappedArray; + hipGetTextureAlignmentOffset; + hipGetTextureReference; + hipUnbindTexture; + hipCreateChannelDesc; + hipCreateTextureObject; + hipDestroyTextureObject; + hipGetChannelDesc; + hipGetTextureObjectResourceDesc; + hipGetTextureObjectResourceViewDesc; + hipGetTextureObjectTextureDesc; + hipTexRefGetAddress; + hipTexRefGetAddressMode; + hipTexRefGetArray; + hipTexRefGetBorderColor; + hipTexRefGetFilterMode; + hipTexRefGetFlags; + hipTexRefGetFormat; + hipTexRefGetMaxAnisotropy; + hipTexRefGetMipmapFilterMode; + hipTexRefGetMipmapLevelBias; + hipTexRefGetMipmapLevelClamp; + hipTexRefGetMipmappedArray; + hipTexRefSetAddress; + hipTexRefSetAddress2D; + hipTexRefSetAddressMode; + hipTexRefSetArray; + hipTexRefSetBorderColor; + hipTexRefSetFilterMode; + hipTexRefSetFlags; + hipTexRefSetFormat; + hipTexRefSetMaxAnisotropy; + hipTexRefSetMipmapFilterMode; + hipTexRefSetMipmapLevelBias; + hipTexRefSetMipmapLevelClamp; + hipTexRefSetMipmappedArray; extern "C++" { hip_impl::hipLaunchKernelGGLImpl*; hip_impl::demangle*; - hipCreateTextureObject*; - hipDestroyTextureObject*; - hipGetTextureObjectResourceDesc*; - hipGetTextureObjectResourceViewDesc*; - hipGetTextureObjectTextureDesc*; - hipBindTexture*; - hipBindTexture2D*; - hipBindTextureToArray*; - hipBindTextureToMipmappedArray*; - hipUnbindTexture*; - hipGetChannelDesc*; - hipGetTextureAlignmentOffset*; - hipGetTextureReference*; - hipTexRefSetFormat*; - hipTexRefSetFlags*; - hipTexRefSetFilterMode*; - hipTexRefSetAddressMode*; - hipTexRefSetArray*; - hipTexRefSetAddress*; - hipTexRefSetAddress2D*; - hipTexRefGetAddress*; - hipTexRefGetAddressMode*; - hipTexRefGetArray*; - hipTexRefSetArray*; - hipCreateChannelDesc*; hipHccGetAccelerator*; hipHccGetAcceleratorView*; hipCreateSurfaceObject*; diff --git a/hipamd/vdi/hip_texture.cpp b/hipamd/vdi/hip_texture.cpp index 6b2f151922..19761d1e78 100644 --- a/hipamd/vdi/hip_texture.cpp +++ b/hipamd/vdi/hip_texture.cpp @@ -671,12 +671,13 @@ hipError_t hipTexRefSetFilterMode(textureReference* texRef, } hipError_t hipTexRefGetAddressMode(hipTextureAddressMode* pam, - textureReference texRef, + const textureReference* texRef, int dim) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetAddressMode, pam, &texRef, dim); + HIP_INIT_API(hipTexRefGetAddressMode, pam, texRef, dim); - if (pam == nullptr) { + if ((pam == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -685,7 +686,7 @@ hipError_t hipTexRefGetAddressMode(hipTextureAddressMode* pam, HIP_RETURN(hipErrorInvalidValue); } - *pam = texRef.addressMode[dim]; + *pam = texRef->addressMode[dim]; HIP_RETURN(hipSuccess); } @@ -709,17 +710,18 @@ hipError_t hipTexRefSetAddressMode(textureReference* texRef, } hipError_t hipTexRefGetArray(hipArray_t* pArray, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetArray, pArray, &texRef); + HIP_INIT_API(hipTexRefGetArray, pArray, texRef); - if (pArray == nullptr) { + if ((pArray == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } hipResourceDesc resDesc = {}; // TODO use ihipGetTextureObjectResourceDesc() to not pollute the API trace. - hipError_t error = hipGetTextureObjectResourceDesc(&resDesc, texRef.textureObject); + hipError_t error = hipGetTextureObjectResourceDesc(&resDesc, texRef->textureObject); if (error != hipSuccess) { return HIP_RETURN(error); } @@ -769,17 +771,18 @@ hipError_t hipTexRefSetArray(textureReference* texRef, } hipError_t hipTexRefGetAddress(hipDeviceptr_t* dptr, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetAddress, dptr, &texRef); + HIP_INIT_API(hipTexRefGetAddress, dptr, texRef); - if (dptr == nullptr) { + if ((dptr == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } hipResourceDesc resDesc = {}; // TODO use ihipGetTextureObjectResourceDesc() to not pollute the API trace. - hipError_t error = hipGetTextureObjectResourceDesc(&resDesc, texRef.textureObject); + hipError_t error = hipGetTextureObjectResourceDesc(&resDesc, texRef->textureObject); if (error != hipSuccess) { return HIP_RETURN(error); } @@ -869,11 +872,12 @@ hipChannelFormatDesc hipCreateChannelDesc(int x, int y, int z, int w, hipChannel } hipError_t hipTexRefGetBorderColor(float* pBorderColor, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetBorderColor, pBorderColor, &texRef); + HIP_INIT_API(hipTexRefGetBorderColor, pBorderColor, texRef); - if (pBorderColor == nullptr) { + if ((pBorderColor == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -885,39 +889,41 @@ hipError_t hipTexRefGetBorderColor(float* pBorderColor, } hipError_t hipTexRefGetFilterMode(hipTextureFilterMode* pfm, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetFilterMode, pfm, &texRef); + HIP_INIT_API(hipTexRefGetFilterMode, pfm, texRef); - if (pfm == nullptr) { + if ((pfm == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } - *pfm = texRef.filterMode; + *pfm = texRef->filterMode; HIP_RETURN(hipSuccess); } hipError_t hipTexRefGetFlags(unsigned int* pFlags, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetFlags, pFlags, &texRef); + HIP_INIT_API(hipTexRefGetFlags, pFlags, texRef); - if (pFlags == nullptr) { + if ((pFlags == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } *pFlags = 0; - if (texRef.readMode == hipReadModeElementType) { + if (texRef->readMode == hipReadModeElementType) { *pFlags |= HIP_TRSF_READ_AS_INTEGER; } - if (texRef.normalized == 1) { + if (texRef->normalized == 1) { *pFlags |= HIP_TRSF_NORMALIZED_COORDINATES; } - if (texRef.sRGB == 1) { + if (texRef->sRGB == 1) { *pFlags |= HIP_TRSF_SRGB; } @@ -926,92 +932,98 @@ hipError_t hipTexRefGetFlags(unsigned int* pFlags, hipError_t hipTexRefGetFormat(hipArray_Format* pFormat, int* pNumChannels, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetFormat, pFormat, pNumChannels, &texRef); + HIP_INIT_API(hipTexRefGetFormat, pFormat, pNumChannels, texRef); if ((pFormat == nullptr) || - (pNumChannels == nullptr)) { + (pNumChannels == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } - *pFormat = texRef.format; - *pNumChannels = texRef.numChannels; + *pFormat = texRef->format; + *pNumChannels = texRef->numChannels; HIP_RETURN(hipSuccess); } hipError_t hipTexRefGetMaxAnisotropy(int* pmaxAnsio, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetMaxAnisotropy, pmaxAnsio, &texRef); + HIP_INIT_API(hipTexRefGetMaxAnisotropy, pmaxAnsio, texRef); - if (pmaxAnsio == nullptr) { + if ((pmaxAnsio == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } - *pmaxAnsio = texRef.maxAnisotropy; + *pmaxAnsio = texRef->maxAnisotropy; HIP_RETURN(hipErrorInvalidValue); } hipError_t hipTexRefGetMipmapFilterMode(hipTextureFilterMode* pfm, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetMipmapFilterMode, pfm, &texRef); + HIP_INIT_API(hipTexRefGetMipmapFilterMode, pfm, texRef); - if (pfm == nullptr) { + if ((pfm == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } - *pfm = texRef.mipmapFilterMode; + *pfm = texRef->mipmapFilterMode; HIP_RETURN(hipErrorInvalidValue); } hipError_t hipTexRefGetMipmapLevelBias(float* pbias, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetMipmapLevelBias, pbias, &texRef); + HIP_INIT_API(hipTexRefGetMipmapLevelBias, pbias, texRef); - if (pbias == nullptr) { + if ((pbias == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } - *pbias = texRef.mipmapLevelBias; + *pbias = texRef->mipmapLevelBias; HIP_RETURN(hipErrorInvalidValue); } hipError_t hipTexRefGetMipmapLevelClamp(float* pminMipmapLevelClamp, float* pmaxMipmapLevelClamp, - textureReference texRef) { + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetMipmapLevelClamp, pminMipmapLevelClamp, pmaxMipmapLevelClamp, &texRef); + HIP_INIT_API(hipTexRefGetMipmapLevelClamp, pminMipmapLevelClamp, pmaxMipmapLevelClamp, texRef); if ((pminMipmapLevelClamp == nullptr) || - (pmaxMipmapLevelClamp == nullptr)){ + (pmaxMipmapLevelClamp == nullptr) || + (texRef == nullptr)){ HIP_RETURN(hipErrorInvalidValue); } - *pminMipmapLevelClamp = texRef.minMipmapLevelClamp; - *pmaxMipmapLevelClamp = texRef.maxMipmapLevelClamp; + *pminMipmapLevelClamp = texRef->minMipmapLevelClamp; + *pmaxMipmapLevelClamp = texRef->maxMipmapLevelClamp; HIP_RETURN(hipErrorInvalidValue); } -hipError_t hipTexRefGetMipMappedArray(hipMipmappedArray_t* pArray, - textureReference texRef) { +hipError_t hipTexRefGetMipmappedArray(hipMipmappedArray_t* pArray, + const textureReference* texRef) { // TODO overload operator<<(ostream&, textureReference&). - HIP_INIT_API(hipTexRefGetMipMappedArray, pArray, &texRef); + HIP_INIT_API(hipTexRefGetMipmappedArray, pArray, &texRef); - if (pArray == nullptr) { + if ((pArray == nullptr) || + (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } hipResourceDesc resDesc = {}; // TODO use ihipGetTextureObjectResourceDesc() to not pollute the API trace. - hipError_t error = hipGetTextureObjectResourceDesc(&resDesc, texRef.textureObject); + hipError_t error = hipGetTextureObjectResourceDesc(&resDesc, texRef->textureObject); if (error != hipSuccess) { return HIP_RETURN(error); }