Reimplement hipMemGetInfo (#1447)

Addresses SWDEV-136570. hipMemGetInfo changed to compute free memory based on information from kfd instead of relying on hc::am_tracker.


[ROCm/hip commit: 3d661e4706]

projects/hip/src/hip_memory.cpp

@@ -19,7 +19,6 @@ LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
 OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
 THE SOFTWARE.
 */
-
 #include <hc_am.hpp>
 #include "hsa/hsa.h"
 #include "hsa/hsa_ext_amd.h"
@@ -28,6 +27,8 @@ THE SOFTWARE.
 #include "hip_hcc_internal.h"
 #include "trace_helper.h"
 
+#include <fstream>
+
 __device__ char __hip_device_heap[__HIP_SIZE_OF_HEAP];
 __device__ uint32_t __hip_device_page_flag[__HIP_NUM_PAGES];
 
@@ -1967,16 +1968,27 @@ hipError_t hipMemGetInfo(size_t* free, size_t* total) {
         } else {
             e = hipErrorInvalidValue;
         }
-
+	
         if (free) {
-            // TODO - replace with kernel-level for reporting free memory:
-            size_t deviceMemSize, hostMemSize, userMemSize;
-            hc::am_memtracker_sizeinfo(device->_acc, &deviceMemSize, &hostMemSize, &userMemSize);
-
-            *free = device->_props.totalGlobalMem - deviceMemSize;
-
-            // Deduct the amount of memory from the free memory reported from the system
-            if (HIP_HIDDEN_FREE_MEM) *free -= (size_t)HIP_HIDDEN_FREE_MEM * 1024 * 1024;
+		if (!device->_driver_node_id) return ihipLogStatus(hipErrorInvalidDevice);
+			
+		std::string fileName = std::string("/sys/class/kfd/kfd/topology/nodes/") + std::to_string(device->_driver_node_id) + std::string("/mem_banks/0/used_memory");  
+		std::ifstream file;
+		file.open(fileName);
+		if (!file) return ihipLogStatus(hipErrorFileNotFound);
+		
+                std::string deviceSize;	
+		size_t deviceMemSize;
+		
+		file >> deviceSize;
+		file.close();                 
+                if ((deviceMemSize=strtol(deviceSize.c_str(),NULL,10))){
+		    *free = device->_props.totalGlobalMem - deviceMemSize;
+		    // Deduct the amount of memory from the free memory reported from the system
+		    if (HIP_HIDDEN_FREE_MEM) *free -= (size_t)HIP_HIDDEN_FREE_MEM * 1024 * 1024;
+		} else {
+ 		    return ihipLogStatus(hipErrorInvalidValue);
+		}
         } else {
             e = hipErrorInvalidValue;
         }