From 7145c500b33d839e9d2d12c6975d7330e2ddbe2f Mon Sep 17 00:00:00 2001
From: kjayapra-amd <karthik.jayaprakash@amd.com>
Date: Mon, 8 Nov 2021 11:05:08 -0500
Subject: [PATCH] SWDEV-298434 - Delete texture references resources before
 deleting the ptr.

Change-Id: Ib616aadd45eced1b10891c908e986bbeddc08288
---
 hipamd/src/hip_global.cpp   |  1 +
 hipamd/src/hip_internal.hpp |  1 +
 hipamd/src/hip_texture.cpp  | 32 +++++++++++++++++++++++---------
 3 files changed, 25 insertions(+), 9 deletions(-)

diff --git a/hipamd/src/hip_global.cpp b/hipamd/src/hip_global.cpp
index c5025d978f..63e30ca51c 100644
--- a/hipamd/src/hip_global.cpp
+++ b/hipamd/src/hip_global.cpp
@@ -41,6 +41,7 @@ DeviceVar::~DeviceVar() {
 
   if (shadowVptr != nullptr) {
     textureReference* texRef = reinterpret_cast<textureReference*>(shadowVptr);
+    ihipUnbindTexture(texRef);
     delete texRef;
     shadowVptr = nullptr;
   }
diff --git a/hipamd/src/hip_internal.hpp b/hipamd/src/hip_internal.hpp
index 40b0b853b7..acbe8145d6 100644
--- a/hipamd/src/hip_internal.hpp
+++ b/hipamd/src/hip_internal.hpp
@@ -371,6 +371,7 @@ extern hipError_t ihipMalloc(void** ptr, size_t sizeBytes, unsigned int flags);
 extern amd::Memory* getMemoryObject(const void* ptr, size_t& offset);
 extern amd::Memory* getMemoryObjectWithOffset(const void* ptr, const size_t size);
 extern void getStreamPerThread(hipStream_t& stream);
+extern hipError_t ihipUnbindTexture(textureReference* texRef);
 
 constexpr bool kOptionChangeable = true;
 constexpr bool kNewDevProg = false;
diff --git a/hipamd/src/hip_texture.cpp b/hipamd/src/hip_texture.cpp
index 7c8995aa25..4ff7f288cd 100644
--- a/hipamd/src/hip_texture.cpp
+++ b/hipamd/src/hip_texture.cpp
@@ -309,7 +309,6 @@ hipError_t hipCreateTextureObject(hipTextureObject_t* pTexObject,
   HIP_RETURN(ihipCreateTextureObject(pTexObject, pResDesc, pTexDesc, pResViewDesc));
 }
 
-
 hipError_t ihipDestroyTextureObject(hipTextureObject_t texObject) {
   if (texObject == nullptr) {
     return hipSuccess;
@@ -332,6 +331,28 @@ hipError_t ihipDestroyTextureObject(hipTextureObject_t texObject) {
   return hipFree(texObject);
 }
 
+hipError_t ihipUnbindTexture(textureReference* texRef) {
+
+  hipError_t hip_error = hipSuccess;
+
+  do {
+    if (texRef == nullptr) {
+      hip_error = hipErrorInvalidValue;
+      break;
+    }
+
+    hip_error = ihipDestroyTextureObject(texRef->textureObject);
+    if (hip_error != hipSuccess) {
+      break;
+    }
+
+    const_cast<textureReference*>(texRef)->textureObject = nullptr;
+
+  } while (0);
+
+  return hip_error;
+}
+
 hipError_t hipDestroyTextureObject(hipTextureObject_t texObject) {
   HIP_INIT_API(hipDestroyTextureObject, texObject);
 
@@ -593,14 +614,7 @@ hipError_t hipBindTextureToMipmappedArray(const textureReference* texref,
 hipError_t hipUnbindTexture(const textureReference* texref) {
   HIP_INIT_API(hipUnbindTexture, texref);
 
-  if (texref == nullptr) {
-    HIP_RETURN(hipErrorInvalidValue);
-  }
-
-  const hipTextureObject_t textureObject = texref->textureObject;
-  const_cast<textureReference*>(texref)->textureObject = nullptr;
-
-  HIP_RETURN(ihipDestroyTextureObject(textureObject));
+  HIP_RETURN(ihipUnbindTexture(const_cast<textureReference*>(texref)));
 }
 
 hipError_t hipBindTexture(size_t* offset,