DEVICE/API: Add NIXL device-side logging infrastructure

src/api/gpu/ucx/nixl_device.cuh

@@ -20,6 +20,27 @@
 #include <nixl_types.h>
 #include <ucp/api/device/ucp_device_impl.h>
 
+/* Helper function to extract the base filename */
+__device__ __forceinline__ static const char* nixl_device_basefile(const char* path) {
+    const char* base = path;
+    for (const char* p = path; *p; ++p) {
+        if (*p == '/') {
+            base = p + 1;
+        }
+    }
+    return base;
+}
+
+/* Helper macro to print a message from NIXL device function including the
+ * thread and block indices, file and line */
+#define nixl_device_printf(_log_level, _fmt, ...) \
+    printf("%c T%-4d:B%-4d%*s%s:%d] " _fmt "\n", _log_level[0], threadIdx.x, blockIdx.x, \
+           17, "", nixl_device_basefile(__FILE__), __LINE__, ##__VA_ARGS__)
+
+/* Print an error message from NIXL device function */
+#define nixl_device_error(_fmt, ...) \
+    nixl_device_printf("ERROR", _fmt, ##__VA_ARGS__)
+
 struct nixlGpuXferStatusH {
     ucp_device_request_t device_request;
 };
@@ -67,7 +88,7 @@ nixlGpuConvertUcsStatus(ucs_status_t status) {
     if (!UCS_STATUS_IS_ERR(status)) {
         return NIXL_IN_PROG;
     }
-    printf("UCX returned error: %d\n", status);
+    nixl_device_error("UCX backend error: %s", ucs_device_status_string(status));
     return NIXL_ERR_BACKEND;
 }
 
@@ -247,6 +268,7 @@ nixlGpuGetXferStatus(nixlGpuXferStatusH &xfer_status) {
     case UCS_INPROGRESS:
         return NIXL_IN_PROG;
     default:
+        nixl_device_error("UCX backend error: %s", ucs_device_status_string(status));
         return NIXL_ERR_BACKEND;
     }
 }