Add hipBindTexture2D on NVCC path (#1773)

hipamd/include/hip/nvcc_detail/hip_runtime_api.h

@@ -1383,6 +1383,12 @@ inline static hipError_t hipBindTexture(size_t* offset, struct textureReference*
     return hipCUDAErrorTohipError(cudaBindTexture(offset, tex, devPtr, desc, size));
 }
 
+inline static hipError_t hipBindTexture2D(size_t* offset, struct textureReference* tex, const void* devPtr,
+                            const hipChannelFormatDesc* desc, size_t width, size_t height,
+                            size_t pitch) {
+    return hipCUDAErrorTohipError(cudaBindTexture2D(offset, tex, devPtr, desc, width, height, pitch));
+}
+
 inline static hipChannelFormatDesc hipCreateChannelDesc(int x, int y, int z, int w,
                                                         hipChannelFormatKind f) {
     return cudaCreateChannelDesc(x, y, z, w, hipChannelFormatKindToCudaChannelFormatKind(f));

hipamd/tests/src/texture/hipBindTex2DPitch.cpp

@@ -58,19 +58,8 @@ int main (void)
             SIZE_W*sizeof(TYPE_t), SIZE_H, hipMemcpyHostToDevice));
 
     tex.normalized = false;
-    #if defined(__HIP_PLATFORM_NVCC__)
-
-    cudaError_t status = cudaBindTexture2D(&tex_ofs, &tex, devPtrA, &tex.channelDesc,
-                                       SIZE_W, SIZE_H, devPitchA);
-    if (status != cudaSuccess) {
-        printf("%serror: '%s'(%d) at %s:%d%s\n", KRED, cudaGetErrorString(status),
-           status,__FILE__, __LINE__, KNRM);
-           failed("API returned error code.");
-    }
-    #else
     HIPCHECK(hipBindTexture2D(&tex_ofs, &tex, devPtrA, &tex.channelDesc,
                                        SIZE_W, SIZE_H, devPitchA));
-    #endif
     HIPCHECK(hipMalloc((void**)&devPtrB, SIZE_W*sizeof(TYPE_t)*SIZE_H)) ;
 
     hipLaunchKernelGGL(texture2dCopyKernel, dim3(4,4,1), dim3(32,32,1), 0, 0, devPtrB);