From 91a7a8d7674acba10edfe1dc23b2d302ddfbd1b9 Mon Sep 17 00:00:00 2001
From: "Manocha, Rahul" <rahul.manocha@amd.com>
Date: Fri, 15 Aug 2025 13:21:52 -0700
Subject: [PATCH] SWDEV-545952 - cuda mappings for StreamGet/Set Attr (#32)

Co-authored-by: Rahul Manocha <rmanocha@amd.com>
---
 .../nvidia_detail/nvidia_hip_runtime_api.h    | 32 +++++++++++++++++++
 1 file changed, 32 insertions(+)

diff --git a/hipnv/include/hip/nvidia_detail/nvidia_hip_runtime_api.h b/hipnv/include/hip/nvidia_detail/nvidia_hip_runtime_api.h
index edf29bdc4d..2be48e3530 100644
--- a/hipnv/include/hip/nvidia_detail/nvidia_hip_runtime_api.h
+++ b/hipnv/include/hip/nvidia_detail/nvidia_hip_runtime_api.h
@@ -28,6 +28,7 @@ THE SOFTWARE.
 #include <cuda_profiler_api.h>
 #include <cuda_fp16.h>
 
+#include <driver_types.h>
 #include <stdio.h>
 
 #define CUDA_9000 9000
@@ -1818,6 +1819,26 @@ typedef enum  cudaKernelNodeAttrID hipKernelNodeAttrID;
 #define hipKernelNodeAttributeCooperative cudaKernelNodeAttributeCooperative
 #define hipKernelNodeAttributePriority cudaKernelNodeAttributePriority
 
+#if CUDA_VERSION >= CUDA_12000
+typedef union cudaStreamAttrValue hipStreamAttrValue;
+typedef enum cudaStreamAttrID hipStreamAttrID;
+#define hipStreamAttributeAccessPolicyWindow cudaStreamAttributeAccessPolicyWindow
+#define hipStreamAttributeSynchronizationPolicy cudaStreamAttributeSynchronizationPolicy
+#define hipStreamAttributeMemSyncDomainMap cudaStreamAttributeMemSyncDomainMap
+#define hipStreamAttributeMemSyncDomain cudaStreamAttributeMemSyncDomain
+#define hipStreamAttributePriority cudaStreamAttributePriority
+
+typedef enum cudaSynchronizationPolicy hipSynchronizationPolicy;
+#define hipSyncPolicyAuto cudaSyncPolicyAuto
+#define hipSyncPolicySpin cudaSyncPolicySpin
+#define hipSyncPolicyYield cudaSyncPolicyYield
+#define hipSyncPolicyBlockingSync cudaSyncPolicyBlockingSync
+
+typedef enum cudaLaunchMemSyncDomain hipLaunchMemSyncDomain;
+#define hipLaunchMemSyncDomainDefault cudaLaunchMemSyncDomainDefault
+#define hipLaunchMemSyncDomainRemote cudaLaunchMemSyncDomainRemote
+#endif
+
 #if CUDA_VERSION >= CUDA_12000
 typedef enum cudaGraphInstantiateResult hipGraphInstantiateResult;
 #define hipGraphInstantiateSuccess cudaGraphInstantiateSuccess
@@ -3215,6 +3236,16 @@ inline static hipError_t hipStreamGetDevice(hipStream_t stream, hipDevice_t* dev
     return hipCUResultTohipError(cuCtxPopCurrent(&context));
 }
 
+inline static hipError_t hipStreamSetAttribute(hipStream_t stream, hipStreamAttrID attr,
+                                               const hipStreamAttrValue* value) {
+    return hipCUDAErrorTohipError(cudaStreamSetAttribute(stream, attr, value));
+}
+
+inline static hipError_t hipStreamGetAttribute(hipStream_t stream, hipStreamAttrID attr,
+                                               hipStreamAttrValue* value_out) {
+    return hipCUDAErrorTohipError(cudaStreamGetAttribute(stream, attr, value_out));
+}
+
 inline static hipError_t hipDriverGetVersion(int* driverVersion) {
     return hipCUDAErrorTohipError(cudaDriverGetVersion(driverVersion));
 }
@@ -4875,6 +4906,7 @@ inline static hipError_t hipSetValidDevices(int* device_arr, int len) {
 
 
 
+
 #endif  //__CUDACC__
 
 #endif  // HIP_INCLUDE_HIP_NVIDIA_DETAIL_HIP_RUNTIME_API_H