diff --git a/projects/clr/hipamd/api/hip/hip_device_runtime.cpp b/projects/clr/hipamd/api/hip/hip_device_runtime.cpp
index fea9fb4dd9..eabf622f28 100644
--- a/projects/clr/hipamd/api/hip/hip_device_runtime.cpp
+++ b/projects/clr/hipamd/api/hip/hip_device_runtime.cpp
@@ -476,3 +476,22 @@ hipError_t hipSetValidDevices ( int* device_arr, int  len ) {
   HIP_RETURN(hipErrorUnknown);
 }
 
+hipError_t hipExtGetLinkTypeAndHopCount(int device1, int device2, uint32_t* linktype, uint32_t* hopcount) {
+  HIP_INIT_API(device1, device2, linktype, hopcount);
+
+  const int numDevices = static_cast<int>(g_devices.size());
+
+  if ((device1 < 0) || (device1 >= numDevices) || (device2 < 0) || (device2 >= numDevices)) {
+    HIP_RETURN(hipErrorInvalidDevice);
+  }
+
+  if (linktype != nullptr) {
+    *linktype = 0;
+  }
+  if (hopcount != nullptr) {
+    *hopcount = 1;
+  }
+
+  HIP_RETURN(hipSuccess);
+}
+
diff --git a/projects/clr/hipamd/api/hip/hip_hcc.def.in b/projects/clr/hipamd/api/hip/hip_hcc.def.in
index 662bfd04c3..f33515ec18 100644
--- a/projects/clr/hipamd/api/hip/hip_hcc.def.in
+++ b/projects/clr/hipamd/api/hip/hip_hcc.def.in
@@ -47,6 +47,8 @@ hipEventElapsedTime
 hipEventQuery
 hipEventRecord
 hipEventSynchronize
+hipExtGetLinkTypeAndHopCount
+hipExtMallocWithFlags
 hipExtModuleLaunchKernel
 hipFree
 hipFreeArray
diff --git a/projects/clr/hipamd/api/hip/hip_hcc.map.in b/projects/clr/hipamd/api/hip/hip_hcc.map.in
index f6ab5a533b..fc8a302b2a 100644
--- a/projects/clr/hipamd/api/hip/hip_hcc.map.in
+++ b/projects/clr/hipamd/api/hip/hip_hcc.map.in
@@ -48,6 +48,8 @@ global:
     hipEventQuery;
     hipEventRecord;
     hipEventSynchronize;
+    hipExtGetLinkTypeAndHopCount;
+    hipExtMallocWithFlags;
     hipExtModuleLaunchKernel;
     hipFree;
     hipFreeArray;
diff --git a/projects/clr/hipamd/api/hip/hip_memory.cpp b/projects/clr/hipamd/api/hip/hip_memory.cpp
index d932c88705..0c0cdcd6ba 100644
--- a/projects/clr/hipamd/api/hip/hip_memory.cpp
+++ b/projects/clr/hipamd/api/hip/hip_memory.cpp
@@ -202,6 +202,17 @@ hipError_t ihipMemset(void* dst, int value, size_t valueSize, size_t sizeBytes,
   return hipSuccess;
 }
 
+hipError_t hipExtMallocWithFlags(void** ptr, size_t sizeBytes, unsigned int flags) {
+  HIP_INIT_API(ptr, sizeBytes, flags);
+
+  if (flags != hipDeviceMallocDefault &&
+      flags != hipDeviceMallocFinegrained) {
+    HIP_RETURN(hipErrorInvalidValue);
+  }
+
+  HIP_RETURN(ihipMalloc(ptr, sizeBytes, (flags & hipDeviceMallocFinegrained)? CL_MEM_SVM_FINE_GRAIN_BUFFER : 0));
+}
+
 hipError_t hipMalloc(void** ptr, size_t sizeBytes) {
   HIP_INIT_API(ptr, sizeBytes);