diff --git a/Data/ThunksDB.json b/Data/ThunksDB.json
index 67f7ef4d2..d52ff0481 100644
--- a/Data/ThunksDB.json
+++ b/Data/ThunksDB.json
@@ -46,6 +46,13 @@
         "@PREFIX_LIB@/libwayland-client.so.0",
         "@PREFIX_LIB@/libwayland-client.so.0.20.0"
       ]
+    },
+    "cuda": {
+      "Library" : "libcuda-guest.so",
+      "Overlay": [
+        "@PREFIX_LIB@/libcuda.so",
+        "@PREFIX_LIB@/libcuda.so.1"
+      ]
     }
   }
 }
diff --git a/ThunkLibs/GuestLibs/CMakeLists.txt b/ThunkLibs/GuestLibs/CMakeLists.txt
index 661971215..0b3069204 100644
--- a/ThunkLibs/GuestLibs/CMakeLists.txt
+++ b/ThunkLibs/GuestLibs/CMakeLists.txt
@@ -266,3 +266,6 @@ add_library(PlaceholderX11 SHARED ../libX11/libX11_NativeGuest.cpp)
 target_link_options(PlaceholderX11 PRIVATE "LINKER:-soname,libX11.so.6")
 set_target_properties(PlaceholderX11 PROPERTIES NO_SONAME ON)
 target_link_libraries(GL-guest PRIVATE PlaceholderX11)
+
+generate(libcuda ${CMAKE_CURRENT_SOURCE_DIR}/../libcuda/libcuda_interface.cpp)
+add_guest_lib(cuda "libcuda.so.1")
diff --git a/ThunkLibs/HostLibs/CMakeLists.txt b/ThunkLibs/HostLibs/CMakeLists.txt
index 28b423420..a526c72ff 100644
--- a/ThunkLibs/HostLibs/CMakeLists.txt
+++ b/ThunkLibs/HostLibs/CMakeLists.txt
@@ -167,6 +167,9 @@ foreach(GUEST_BITNESS IN LISTS BITNESS_LIST)
 
   find_package(OpenGL REQUIRED)
   target_link_libraries(GL-host-${GUEST_BITNESS} PRIVATE OpenGL::GL)
+
+  generate(libcuda ${CMAKE_CURRENT_SOURCE_DIR}/../libcuda/libcuda_interface.cpp ${GUEST_BITNESS})
+  add_host_lib(cuda ${GUEST_BITNESS})
 endforeach()
 
 if (BUILD_FEX_LINUX_TESTS)
diff --git a/ThunkLibs/libcuda/cuda_defines.h b/ThunkLibs/libcuda/cuda_defines.h
new file mode 100644
index 000000000..7a0522156
--- /dev/null
+++ b/ThunkLibs/libcuda/cuda_defines.h
@@ -0,0 +1,1017 @@
+// SPDX-License-Identifier: MIT
+// Generated file. Be careful if modifying
+#pragma once
+#include <stdint.h>
+#include "fixup_defines.h"
+#ifdef __cplusplus
+extern "C" {
+#endif
+
+// Type forward declarations
+typedef uint32_t cuuint32_t;
+typedef uint64_t cuuint64_t;
+typedef unsigned long long CUdeviceptr_v2;
+typedef CUdeviceptr_v2 CUdeviceptr;
+typedef int CUdevice_v1;
+typedef CUdevice_v1 CUdevice;
+struct CUctx_st;
+typedef struct CUctx_st* CUcontext;
+struct CUmod_st;
+typedef struct CUmod_st* CUmodule;
+struct CUfunc_st;
+typedef struct CUfunc_st* CUfunction;
+struct CUlib_st;
+typedef struct CUlib_st* CUlibrary;
+struct CUkern_st;
+typedef struct CUkern_st* CUkernel;
+struct CUarray_st;
+typedef struct CUarray_st* CUarray;
+struct CUmipmappedArray_st;
+typedef struct CUmipmappedArray_st* CUmipmappedArray;
+struct CUtexref_st;
+typedef struct CUtexref_st* CUtexref;
+struct CUsurfref_st;
+typedef struct CUsurfref_st* CUsurfref;
+struct CUevent_st;
+typedef struct CUevent_st* CUevent;
+struct CUstream_st;
+typedef struct CUstream_st* CUstream;
+struct CUgraphicsResource_st;
+typedef struct CUgraphicsResource_st* CUgraphicsResource;
+typedef unsigned long long CUtexObject_v1;
+typedef CUtexObject_v1 CUtexObject;
+typedef unsigned long long CUsurfObject_v1;
+typedef CUsurfObject_v1 CUsurfObject;
+struct CUextMemory_st;
+typedef struct CUextMemory_st* CUexternalMemory;
+struct CUextSemaphore_st;
+typedef struct CUextSemaphore_st* CUexternalSemaphore;
+struct CUgraph_st;
+typedef struct CUgraph_st* CUgraph;
+struct CUgraphNode_st;
+typedef struct CUgraphNode_st* CUgraphNode;
+struct CUgraphExec_st;
+typedef struct CUgraphExec_st* CUgraphExec;
+struct CUmemPoolHandle_st;
+typedef struct CUmemPoolHandle_st* CUmemoryPool;
+struct CUuserObject_st;
+typedef struct CUuserObject_st* CUuserObject;
+typedef cuuint64_t CUgraphConditionalHandle;
+struct CUgraphDeviceUpdatableNode_st;
+typedef struct CUgraphDeviceUpdatableNode_st* CUgraphDeviceNode;
+struct CUasyncCallbackEntry_st;
+typedef struct CUasyncCallbackEntry_st* CUasyncCallbackHandle;
+struct CUuuid_st {
+  char _0[16];
+};
+typedef struct CUuuid_st CUuuid;
+struct CUmemFabricHandle_st {
+  unsigned char _0[64];
+};
+typedef struct CUmemFabricHandle_st CUmemFabricHandle_v1;
+typedef CUmemFabricHandle_v1 CUmemFabricHandle;
+struct CUipcEventHandle_st {
+  char _0[64];
+};
+typedef struct CUipcEventHandle_st CUipcEventHandle_v1;
+typedef CUipcEventHandle_v1 CUipcEventHandle;
+struct CUipcMemHandle_st {
+  char _0[64];
+};
+typedef struct CUipcMemHandle_st CUipcMemHandle_v1;
+typedef CUipcMemHandle_v1 CUipcMemHandle;
+enum CUipcMem_flags_enum : unsigned int {};
+typedef enum CUipcMem_flags_enum CUipcMem_flags;
+enum CUmemAttach_flags_enum : unsigned int {};
+typedef enum CUmemAttach_flags_enum CUmemAttach_flags;
+enum CUctx_flags_enum : unsigned int {};
+typedef enum CUctx_flags_enum CUctx_flags;
+enum CUevent_sched_flags_enum : unsigned int {};
+typedef enum CUevent_sched_flags_enum CUevent_sched_flags;
+enum CUstream_flags_enum : unsigned int {};
+typedef enum CUstream_flags_enum CUstream_flags;
+enum CUevent_flags_enum : unsigned int {};
+typedef enum CUevent_flags_enum CUevent_flags;
+enum CUevent_record_flags_enum : unsigned int {};
+typedef enum CUevent_record_flags_enum CUevent_record_flags;
+enum CUevent_wait_flags_enum : unsigned int {};
+typedef enum CUevent_wait_flags_enum CUevent_wait_flags;
+enum CUstreamWaitValue_flags_enum : unsigned int {};
+typedef enum CUstreamWaitValue_flags_enum CUstreamWaitValue_flags;
+enum CUstreamWriteValue_flags_enum : unsigned int {};
+typedef enum CUstreamWriteValue_flags_enum CUstreamWriteValue_flags;
+enum CUstreamBatchMemOpType_enum : unsigned int {};
+typedef enum CUstreamBatchMemOpType_enum CUstreamBatchMemOpType;
+enum CUstreamMemoryBarrier_flags_enum : unsigned int {};
+typedef enum CUstreamMemoryBarrier_flags_enum CUstreamMemoryBarrier_flags;
+union CUstreamBatchMemOpParams_union;
+typedef union CUstreamBatchMemOpParams_union CUstreamBatchMemOpParams_v1;
+typedef CUstreamBatchMemOpParams_v1 CUstreamBatchMemOpParams;
+struct CUDA_BATCH_MEM_OP_NODE_PARAMS_v1_st;
+typedef struct CUDA_BATCH_MEM_OP_NODE_PARAMS_v1_st CUDA_BATCH_MEM_OP_NODE_PARAMS_v1;
+typedef CUDA_BATCH_MEM_OP_NODE_PARAMS_v1 CUDA_BATCH_MEM_OP_NODE_PARAMS;
+struct CUDA_BATCH_MEM_OP_NODE_PARAMS_v2_st;
+typedef struct CUDA_BATCH_MEM_OP_NODE_PARAMS_v2_st CUDA_BATCH_MEM_OP_NODE_PARAMS_v2;
+enum CUoccupancy_flags_enum : unsigned int {};
+typedef enum CUoccupancy_flags_enum CUoccupancy_flags;
+enum CUstreamUpdateCaptureDependencies_flags_enum : unsigned int {};
+typedef enum CUstreamUpdateCaptureDependencies_flags_enum CUstreamUpdateCaptureDependencies_flags;
+enum CUasyncNotificationType_enum : unsigned int {};
+typedef enum CUasyncNotificationType_enum CUasyncNotificationType;
+struct CUasyncNotificationInfo_st;
+typedef struct CUasyncNotificationInfo_st CUasyncNotificationInfo;
+enum CUarray_format_enum : unsigned int {};
+typedef enum CUarray_format_enum CUarray_format;
+enum CUaddress_mode_enum : unsigned int {};
+typedef enum CUaddress_mode_enum CUaddress_mode;
+enum CUfilter_mode_enum : unsigned int {};
+typedef enum CUfilter_mode_enum CUfilter_mode;
+enum CUdevice_attribute_enum : unsigned int {};
+typedef enum CUdevice_attribute_enum CUdevice_attribute;
+struct CUdevprop_st {
+  int _0;
+  int _1[3];
+  int _2[3];
+  int _3;
+  int _4;
+  int _5;
+  int _6;
+  int _7;
+  int _8;
+  int _9;
+};
+typedef struct CUdevprop_st CUdevprop_v1;
+typedef CUdevprop_v1 CUdevprop;
+enum CUpointer_attribute_enum : unsigned int {};
+typedef enum CUpointer_attribute_enum CUpointer_attribute;
+enum CUfunction_attribute_enum : unsigned int {};
+typedef enum CUfunction_attribute_enum CUfunction_attribute;
+enum CUfunc_cache_enum : unsigned int {};
+typedef enum CUfunc_cache_enum CUfunc_cache;
+enum CUsharedconfig_enum : unsigned int {};
+typedef enum CUsharedconfig_enum CUsharedconfig;
+enum CUshared_carveout_enum : int {};
+typedef enum CUshared_carveout_enum CUshared_carveout;
+enum CUmemorytype_enum : unsigned int {};
+typedef enum CUmemorytype_enum CUmemorytype;
+enum CUcomputemode_enum : unsigned int {};
+typedef enum CUcomputemode_enum CUcomputemode;
+enum CUmem_advise_enum : unsigned int {};
+typedef enum CUmem_advise_enum CUmem_advise;
+enum CUmem_range_attribute_enum : unsigned int {};
+typedef enum CUmem_range_attribute_enum CUmem_range_attribute;
+enum CUjit_option_enum : unsigned int {};
+typedef enum CUjit_option_enum CUjit_option;
+enum CUjit_target_enum : unsigned int {};
+typedef enum CUjit_target_enum CUjit_target;
+enum CUjit_fallback_enum : unsigned int {};
+typedef enum CUjit_fallback_enum CUjit_fallback;
+enum CUjit_cacheMode_enum : unsigned int {};
+typedef enum CUjit_cacheMode_enum CUjit_cacheMode;
+enum CUjitInputType_enum : unsigned int {};
+typedef enum CUjitInputType_enum CUjitInputType;
+struct CUlinkState_st;
+typedef struct CUlinkState_st* CUlinkState;
+enum CUgraphicsRegisterFlags_enum : unsigned int {};
+typedef enum CUgraphicsRegisterFlags_enum CUgraphicsRegisterFlags;
+enum CUgraphicsMapResourceFlags_enum : unsigned int {};
+typedef enum CUgraphicsMapResourceFlags_enum CUgraphicsMapResourceFlags;
+enum CUarray_cubemap_face_enum : unsigned int {};
+typedef enum CUarray_cubemap_face_enum CUarray_cubemap_face;
+enum CUlimit_enum : unsigned int {};
+typedef enum CUlimit_enum CUlimit;
+enum CUresourcetype_enum : unsigned int {};
+typedef enum CUresourcetype_enum CUresourcetype;
+enum CUaccessProperty_enum : unsigned int {};
+typedef enum CUaccessProperty_enum CUaccessProperty;
+struct CUaccessPolicyWindow_st;
+typedef struct CUaccessPolicyWindow_st CUaccessPolicyWindow_v1;
+typedef CUaccessPolicyWindow_v1 CUaccessPolicyWindow;
+struct CUDA_KERNEL_NODE_PARAMS_st;
+typedef struct CUDA_KERNEL_NODE_PARAMS_st CUDA_KERNEL_NODE_PARAMS_v1;
+struct CUDA_KERNEL_NODE_PARAMS_v2_st;
+typedef struct CUDA_KERNEL_NODE_PARAMS_v2_st CUDA_KERNEL_NODE_PARAMS_v2;
+typedef CUDA_KERNEL_NODE_PARAMS_v2 CUDA_KERNEL_NODE_PARAMS;
+struct CUDA_KERNEL_NODE_PARAMS_v3_st;
+typedef struct CUDA_KERNEL_NODE_PARAMS_v3_st CUDA_KERNEL_NODE_PARAMS_v3;
+struct CUDA_MEMSET_NODE_PARAMS_st;
+typedef struct CUDA_MEMSET_NODE_PARAMS_st CUDA_MEMSET_NODE_PARAMS_v1;
+typedef CUDA_MEMSET_NODE_PARAMS_v1 CUDA_MEMSET_NODE_PARAMS;
+struct CUDA_MEMSET_NODE_PARAMS_v2_st;
+typedef struct CUDA_MEMSET_NODE_PARAMS_v2_st CUDA_MEMSET_NODE_PARAMS_v2;
+struct CUDA_HOST_NODE_PARAMS_st {
+  CUhostFn _0;
+  void* _1;
+};
+typedef struct CUDA_HOST_NODE_PARAMS_st CUDA_HOST_NODE_PARAMS_v1;
+typedef CUDA_HOST_NODE_PARAMS_v1 CUDA_HOST_NODE_PARAMS;
+struct CUDA_HOST_NODE_PARAMS_v2_st {
+  CUhostFn _0;
+  void* _1;
+};
+typedef struct CUDA_HOST_NODE_PARAMS_v2_st CUDA_HOST_NODE_PARAMS_v2;
+enum CUgraphConditionalNodeType_enum : unsigned int {};
+typedef enum CUgraphConditionalNodeType_enum CUgraphConditionalNodeType;
+enum CUgraphNodeType_enum : unsigned int {};
+typedef enum CUgraphNodeType_enum CUgraphNodeType;
+enum CUgraphDependencyType_enum : unsigned int {};
+typedef enum CUgraphDependencyType_enum CUgraphDependencyType;
+struct CUgraphEdgeData_st {
+  unsigned char _0;
+  unsigned char _1;
+  unsigned char _2;
+  unsigned char _3[5];
+};
+typedef struct CUgraphEdgeData_st CUgraphEdgeData;
+enum CUgraphInstantiateResult_enum : unsigned int {};
+typedef enum CUgraphInstantiateResult_enum CUgraphInstantiateResult;
+struct CUDA_GRAPH_INSTANTIATE_PARAMS_st;
+typedef struct CUDA_GRAPH_INSTANTIATE_PARAMS_st CUDA_GRAPH_INSTANTIATE_PARAMS;
+enum CUsynchronizationPolicy_enum : unsigned int {};
+typedef enum CUsynchronizationPolicy_enum CUsynchronizationPolicy;
+enum CUclusterSchedulingPolicy_enum : unsigned int {};
+typedef enum CUclusterSchedulingPolicy_enum CUclusterSchedulingPolicy;
+enum CUlaunchMemSyncDomain_enum : unsigned int {};
+typedef enum CUlaunchMemSyncDomain_enum CUlaunchMemSyncDomain;
+struct CUlaunchMemSyncDomainMap_st {
+  unsigned char _0;
+  unsigned char _1;
+};
+typedef struct CUlaunchMemSyncDomainMap_st CUlaunchMemSyncDomainMap;
+enum CUlaunchAttributeID_enum : unsigned int {};
+typedef enum CUlaunchAttributeID_enum CUlaunchAttributeID;
+union CUlaunchAttributeValue_union;
+typedef union CUlaunchAttributeValue_union CUlaunchAttributeValue;
+struct CUlaunchAttribute_st;
+typedef struct CUlaunchAttribute_st CUlaunchAttribute;
+struct CUlaunchConfig_st;
+typedef struct CUlaunchConfig_st CUlaunchConfig;
+typedef CUlaunchAttributeID CUkernelNodeAttrID;
+typedef CUlaunchAttributeValue CUkernelNodeAttrValue_v1;
+typedef CUkernelNodeAttrValue_v1 CUkernelNodeAttrValue;
+enum CUstreamCaptureStatus_enum : unsigned int {};
+typedef enum CUstreamCaptureStatus_enum CUstreamCaptureStatus;
+enum CUstreamCaptureMode_enum : unsigned int {};
+typedef enum CUstreamCaptureMode_enum CUstreamCaptureMode;
+typedef CUlaunchAttributeID CUstreamAttrID;
+typedef CUlaunchAttributeValue CUstreamAttrValue_v1;
+typedef CUstreamAttrValue_v1 CUstreamAttrValue;
+enum CUdriverProcAddress_flags_enum : unsigned int {};
+typedef enum CUdriverProcAddress_flags_enum CUdriverProcAddress_flags;
+enum CUdriverProcAddressQueryResult_enum : unsigned int {};
+typedef enum CUdriverProcAddressQueryResult_enum CUdriverProcAddressQueryResult;
+enum CUexecAffinityType_enum : unsigned int {};
+typedef enum CUexecAffinityType_enum CUexecAffinityType;
+struct CUexecAffinitySmCount_st {
+  unsigned int _0;
+};
+typedef struct CUexecAffinitySmCount_st CUexecAffinitySmCount_v1;
+typedef CUexecAffinitySmCount_v1 CUexecAffinitySmCount;
+struct CUexecAffinityParam_st;
+typedef struct CUexecAffinityParam_st CUexecAffinityParam_v1;
+typedef CUexecAffinityParam_v1 CUexecAffinityParam;
+enum CUlibraryOption_enum : unsigned int {};
+typedef enum CUlibraryOption_enum CUlibraryOption;
+struct CUlibraryHostUniversalFunctionAndDataTable_st {
+  void* _0;
+  size_t _1;
+  void* _2;
+  size_t _3;
+};
+typedef struct CUlibraryHostUniversalFunctionAndDataTable_st CUlibraryHostUniversalFunctionAndDataTable;
+typedef enum cudaError_enum CUresult;
+enum CUdevice_P2PAttribute_enum : unsigned int {};
+typedef enum CUdevice_P2PAttribute_enum CUdevice_P2PAttribute;
+struct CUDA_MEMCPY2D_st;
+typedef struct CUDA_MEMCPY2D_st CUDA_MEMCPY2D_v2;
+typedef CUDA_MEMCPY2D_v2 CUDA_MEMCPY2D;
+struct CUDA_MEMCPY3D_st {
+  size_t _0;
+  size_t _1;
+  size_t _2;
+  size_t _3;
+  CUmemorytype _4;
+  const void* _5;
+  CUdeviceptr _6;
+  CUarray _7;
+  void* _8;
+  size_t _9;
+  size_t _10;
+  size_t _11;
+  size_t _12;
+  size_t _13;
+  size_t _14;
+  CUmemorytype _15;
+  void* _16;
+  CUdeviceptr _17;
+  CUarray _18;
+  void* _19;
+  size_t _20;
+  size_t _21;
+  size_t _22;
+  size_t _23;
+  size_t _24;
+};
+typedef struct CUDA_MEMCPY3D_st CUDA_MEMCPY3D_v2;
+typedef CUDA_MEMCPY3D_v2 CUDA_MEMCPY3D;
+struct CUDA_MEMCPY3D_PEER_st {
+  size_t _0;
+  size_t _1;
+  size_t _2;
+  size_t _3;
+  CUmemorytype _4;
+  const void* _5;
+  CUdeviceptr _6;
+  CUarray _7;
+  CUcontext _8;
+  size_t _9;
+  size_t _10;
+  size_t _11;
+  size_t _12;
+  size_t _13;
+  size_t _14;
+  CUmemorytype _15;
+  void* _16;
+  CUdeviceptr _17;
+  CUarray _18;
+  CUcontext _19;
+  size_t _20;
+  size_t _21;
+  size_t _22;
+  size_t _23;
+  size_t _24;
+};
+typedef struct CUDA_MEMCPY3D_PEER_st CUDA_MEMCPY3D_PEER_v1;
+typedef CUDA_MEMCPY3D_PEER_v1 CUDA_MEMCPY3D_PEER;
+struct CUDA_MEMCPY_NODE_PARAMS_st;
+typedef struct CUDA_MEMCPY_NODE_PARAMS_st CUDA_MEMCPY_NODE_PARAMS;
+struct CUDA_ARRAY_DESCRIPTOR_st;
+typedef struct CUDA_ARRAY_DESCRIPTOR_st CUDA_ARRAY_DESCRIPTOR_v2;
+typedef CUDA_ARRAY_DESCRIPTOR_v2 CUDA_ARRAY_DESCRIPTOR;
+struct CUDA_ARRAY3D_DESCRIPTOR_st;
+typedef struct CUDA_ARRAY3D_DESCRIPTOR_st CUDA_ARRAY3D_DESCRIPTOR_v2;
+typedef CUDA_ARRAY3D_DESCRIPTOR_v2 CUDA_ARRAY3D_DESCRIPTOR;
+struct CUDA_ARRAY_SPARSE_PROPERTIES_st;
+typedef struct CUDA_ARRAY_SPARSE_PROPERTIES_st CUDA_ARRAY_SPARSE_PROPERTIES_v1;
+typedef CUDA_ARRAY_SPARSE_PROPERTIES_v1 CUDA_ARRAY_SPARSE_PROPERTIES;
+struct CUDA_ARRAY_MEMORY_REQUIREMENTS_st {
+  size_t _0;
+  size_t _1;
+  unsigned int _2[4];
+};
+typedef struct CUDA_ARRAY_MEMORY_REQUIREMENTS_st CUDA_ARRAY_MEMORY_REQUIREMENTS_v1;
+typedef CUDA_ARRAY_MEMORY_REQUIREMENTS_v1 CUDA_ARRAY_MEMORY_REQUIREMENTS;
+struct CUDA_RESOURCE_DESC_st;
+typedef struct CUDA_RESOURCE_DESC_st CUDA_RESOURCE_DESC_v1;
+typedef CUDA_RESOURCE_DESC_v1 CUDA_RESOURCE_DESC;
+struct CUDA_TEXTURE_DESC_st;
+typedef struct CUDA_TEXTURE_DESC_st CUDA_TEXTURE_DESC_v1;
+typedef CUDA_TEXTURE_DESC_v1 CUDA_TEXTURE_DESC;
+enum CUresourceViewFormat_enum : unsigned int {};
+typedef enum CUresourceViewFormat_enum CUresourceViewFormat;
+struct CUDA_RESOURCE_VIEW_DESC_st;
+typedef struct CUDA_RESOURCE_VIEW_DESC_st CUDA_RESOURCE_VIEW_DESC_v1;
+typedef CUDA_RESOURCE_VIEW_DESC_v1 CUDA_RESOURCE_VIEW_DESC;
+struct CUtensorMap_st {
+  cuuint64_t _0[16];
+};
+typedef struct CUtensorMap_st CUtensorMap;
+enum CUtensorMapDataType_enum : unsigned int {};
+typedef enum CUtensorMapDataType_enum CUtensorMapDataType;
+enum CUtensorMapInterleave_enum : unsigned int {};
+typedef enum CUtensorMapInterleave_enum CUtensorMapInterleave;
+enum CUtensorMapSwizzle_enum : unsigned int {};
+typedef enum CUtensorMapSwizzle_enum CUtensorMapSwizzle;
+enum CUtensorMapL2promotion_enum : unsigned int {};
+typedef enum CUtensorMapL2promotion_enum CUtensorMapL2promotion;
+enum CUtensorMapFloatOOBfill_enum : unsigned int {};
+typedef enum CUtensorMapFloatOOBfill_enum CUtensorMapFloatOOBfill;
+struct CUDA_POINTER_ATTRIBUTE_P2P_TOKENS_st {
+  unsigned long long _0;
+  unsigned int _1;
+};
+typedef struct CUDA_POINTER_ATTRIBUTE_P2P_TOKENS_st CUDA_POINTER_ATTRIBUTE_P2P_TOKENS_v1;
+typedef CUDA_POINTER_ATTRIBUTE_P2P_TOKENS_v1 CUDA_POINTER_ATTRIBUTE_P2P_TOKENS;
+enum CUDA_POINTER_ATTRIBUTE_ACCESS_FLAGS_enum : unsigned int {};
+typedef enum CUDA_POINTER_ATTRIBUTE_ACCESS_FLAGS_enum CUDA_POINTER_ATTRIBUTE_ACCESS_FLAGS;
+struct CUDA_LAUNCH_PARAMS_st {
+  CUfunction _0;
+  unsigned int _1;
+  unsigned int _2;
+  unsigned int _3;
+  unsigned int _4;
+  unsigned int _5;
+  unsigned int _6;
+  unsigned int _7;
+  CUstream _8;
+  void** _9;
+};
+typedef struct CUDA_LAUNCH_PARAMS_st CUDA_LAUNCH_PARAMS_v1;
+typedef CUDA_LAUNCH_PARAMS_v1 CUDA_LAUNCH_PARAMS;
+enum CUexternalMemoryHandleType_enum : unsigned int {};
+typedef enum CUexternalMemoryHandleType_enum CUexternalMemoryHandleType;
+struct CUDA_EXTERNAL_MEMORY_HANDLE_DESC_st;
+typedef struct CUDA_EXTERNAL_MEMORY_HANDLE_DESC_st CUDA_EXTERNAL_MEMORY_HANDLE_DESC_v1;
+typedef CUDA_EXTERNAL_MEMORY_HANDLE_DESC_v1 CUDA_EXTERNAL_MEMORY_HANDLE_DESC;
+struct CUDA_EXTERNAL_MEMORY_BUFFER_DESC_st {
+  unsigned long long _0;
+  unsigned long long _1;
+  unsigned int _2;
+  unsigned int _3[16];
+};
+typedef struct CUDA_EXTERNAL_MEMORY_BUFFER_DESC_st CUDA_EXTERNAL_MEMORY_BUFFER_DESC_v1;
+typedef CUDA_EXTERNAL_MEMORY_BUFFER_DESC_v1 CUDA_EXTERNAL_MEMORY_BUFFER_DESC;
+struct CUDA_EXTERNAL_MEMORY_MIPMAPPED_ARRAY_DESC_st;
+typedef struct CUDA_EXTERNAL_MEMORY_MIPMAPPED_ARRAY_DESC_st CUDA_EXTERNAL_MEMORY_MIPMAPPED_ARRAY_DESC_v1;
+typedef CUDA_EXTERNAL_MEMORY_MIPMAPPED_ARRAY_DESC_v1 CUDA_EXTERNAL_MEMORY_MIPMAPPED_ARRAY_DESC;
+enum CUexternalSemaphoreHandleType_enum : unsigned int {};
+typedef enum CUexternalSemaphoreHandleType_enum CUexternalSemaphoreHandleType;
+struct CUDA_EXTERNAL_SEMAPHORE_HANDLE_DESC_st;
+typedef struct CUDA_EXTERNAL_SEMAPHORE_HANDLE_DESC_st CUDA_EXTERNAL_SEMAPHORE_HANDLE_DESC_v1;
+typedef CUDA_EXTERNAL_SEMAPHORE_HANDLE_DESC_v1 CUDA_EXTERNAL_SEMAPHORE_HANDLE_DESC;
+struct CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS_st;
+typedef struct CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS_st CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS_v1;
+typedef CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS_v1 CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS;
+struct CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS_st;
+typedef struct CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS_st CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS_v1;
+typedef CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS_v1 CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS;
+struct CUDA_EXT_SEM_SIGNAL_NODE_PARAMS_st {
+  CUexternalSemaphore* _0;
+  const CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS* _1;
+  unsigned int _2;
+};
+typedef struct CUDA_EXT_SEM_SIGNAL_NODE_PARAMS_st CUDA_EXT_SEM_SIGNAL_NODE_PARAMS_v1;
+typedef CUDA_EXT_SEM_SIGNAL_NODE_PARAMS_v1 CUDA_EXT_SEM_SIGNAL_NODE_PARAMS;
+struct CUDA_EXT_SEM_SIGNAL_NODE_PARAMS_v2_st {
+  CUexternalSemaphore* _0;
+  const CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS* _1;
+  unsigned int _2;
+};
+typedef struct CUDA_EXT_SEM_SIGNAL_NODE_PARAMS_v2_st CUDA_EXT_SEM_SIGNAL_NODE_PARAMS_v2;
+struct CUDA_EXT_SEM_WAIT_NODE_PARAMS_st {
+  CUexternalSemaphore* _0;
+  const CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS* _1;
+  unsigned int _2;
+};
+typedef struct CUDA_EXT_SEM_WAIT_NODE_PARAMS_st CUDA_EXT_SEM_WAIT_NODE_PARAMS_v1;
+typedef CUDA_EXT_SEM_WAIT_NODE_PARAMS_v1 CUDA_EXT_SEM_WAIT_NODE_PARAMS;
+struct CUDA_EXT_SEM_WAIT_NODE_PARAMS_v2_st {
+  CUexternalSemaphore* _0;
+  const CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS* _1;
+  unsigned int _2;
+};
+typedef struct CUDA_EXT_SEM_WAIT_NODE_PARAMS_v2_st CUDA_EXT_SEM_WAIT_NODE_PARAMS_v2;
+typedef unsigned long long CUmemGenericAllocationHandle_v1;
+typedef CUmemGenericAllocationHandle_v1 CUmemGenericAllocationHandle;
+enum CUmemAllocationHandleType_enum : unsigned int {};
+typedef enum CUmemAllocationHandleType_enum CUmemAllocationHandleType;
+enum CUmemAccess_flags_enum : unsigned int {};
+typedef enum CUmemAccess_flags_enum CUmemAccess_flags;
+enum CUmemLocationType_enum : unsigned int {};
+typedef enum CUmemLocationType_enum CUmemLocationType;
+enum CUmemAllocationType_enum : unsigned int {};
+typedef enum CUmemAllocationType_enum CUmemAllocationType;
+enum CUmemAllocationGranularity_flags_enum : unsigned int {};
+typedef enum CUmemAllocationGranularity_flags_enum CUmemAllocationGranularity_flags;
+enum CUmemRangeHandleType_enum : unsigned int {};
+typedef enum CUmemRangeHandleType_enum CUmemRangeHandleType;
+enum CUarraySparseSubresourceType_enum : unsigned int {};
+typedef enum CUarraySparseSubresourceType_enum CUarraySparseSubresourceType;
+enum CUmemOperationType_enum : unsigned int {};
+typedef enum CUmemOperationType_enum CUmemOperationType;
+enum CUmemHandleType_enum : unsigned int {};
+typedef enum CUmemHandleType_enum CUmemHandleType;
+struct CUarrayMapInfo_st;
+typedef struct CUarrayMapInfo_st CUarrayMapInfo_v1;
+typedef CUarrayMapInfo_v1 CUarrayMapInfo;
+struct CUmemLocation_st {
+  CUmemLocationType _0;
+  int _1;
+};
+typedef struct CUmemLocation_st CUmemLocation_v1;
+typedef CUmemLocation_v1 CUmemLocation;
+enum CUmemAllocationCompType_enum : unsigned int {};
+typedef enum CUmemAllocationCompType_enum CUmemAllocationCompType;
+struct CUmemAllocationProp_st;
+typedef struct CUmemAllocationProp_st CUmemAllocationProp_v1;
+typedef CUmemAllocationProp_v1 CUmemAllocationProp;
+enum CUmulticastGranularity_flags_enum : unsigned int {};
+typedef enum CUmulticastGranularity_flags_enum CUmulticastGranularity_flags;
+struct CUmulticastObjectProp_st {
+  unsigned int _0;
+  size_t _1;
+  unsigned long long _2;
+  unsigned long long _3;
+};
+typedef struct CUmulticastObjectProp_st CUmulticastObjectProp_v1;
+typedef CUmulticastObjectProp_v1 CUmulticastObjectProp;
+struct CUmemAccessDesc_st;
+typedef struct CUmemAccessDesc_st CUmemAccessDesc_v1;
+typedef CUmemAccessDesc_v1 CUmemAccessDesc;
+enum CUgraphExecUpdateResult_enum : unsigned int {};
+typedef enum CUgraphExecUpdateResult_enum CUgraphExecUpdateResult;
+struct CUgraphExecUpdateResultInfo_st;
+typedef struct CUgraphExecUpdateResultInfo_st CUgraphExecUpdateResultInfo_v1;
+typedef CUgraphExecUpdateResultInfo_v1 CUgraphExecUpdateResultInfo;
+enum CUmemPool_attribute_enum : unsigned int {};
+typedef enum CUmemPool_attribute_enum CUmemPool_attribute;
+struct CUmemPoolProps_st;
+typedef struct CUmemPoolProps_st CUmemPoolProps_v1;
+typedef CUmemPoolProps_v1 CUmemPoolProps;
+struct CUmemPoolPtrExportData_st {
+  unsigned char _0[64];
+};
+typedef struct CUmemPoolPtrExportData_st CUmemPoolPtrExportData_v1;
+typedef CUmemPoolPtrExportData_v1 CUmemPoolPtrExportData;
+struct CUDA_MEM_ALLOC_NODE_PARAMS_v1_st;
+typedef struct CUDA_MEM_ALLOC_NODE_PARAMS_v1_st CUDA_MEM_ALLOC_NODE_PARAMS_v1;
+typedef CUDA_MEM_ALLOC_NODE_PARAMS_v1 CUDA_MEM_ALLOC_NODE_PARAMS;
+struct CUDA_MEM_ALLOC_NODE_PARAMS_v2_st;
+typedef struct CUDA_MEM_ALLOC_NODE_PARAMS_v2_st CUDA_MEM_ALLOC_NODE_PARAMS_v2;
+struct CUDA_MEM_FREE_NODE_PARAMS_st;
+typedef struct CUDA_MEM_FREE_NODE_PARAMS_st CUDA_MEM_FREE_NODE_PARAMS;
+enum CUgraphMem_attribute_enum : unsigned int {};
+typedef enum CUgraphMem_attribute_enum CUgraphMem_attribute;
+struct CUDA_CHILD_GRAPH_NODE_PARAMS_st;
+typedef struct CUDA_CHILD_GRAPH_NODE_PARAMS_st CUDA_CHILD_GRAPH_NODE_PARAMS;
+struct CUDA_EVENT_RECORD_NODE_PARAMS_st;
+typedef struct CUDA_EVENT_RECORD_NODE_PARAMS_st CUDA_EVENT_RECORD_NODE_PARAMS;
+struct CUDA_EVENT_WAIT_NODE_PARAMS_st;
+typedef struct CUDA_EVENT_WAIT_NODE_PARAMS_st CUDA_EVENT_WAIT_NODE_PARAMS;
+struct CUgraphNodeParams_st;
+typedef struct CUgraphNodeParams_st CUgraphNodeParams;
+enum CUflushGPUDirectRDMAWritesOptions_enum : unsigned int {};
+typedef enum CUflushGPUDirectRDMAWritesOptions_enum CUflushGPUDirectRDMAWritesOptions;
+enum CUGPUDirectRDMAWritesOrdering_enum : unsigned int {};
+typedef enum CUGPUDirectRDMAWritesOrdering_enum CUGPUDirectRDMAWritesOrdering;
+enum CUflushGPUDirectRDMAWritesScope_enum : unsigned int {};
+typedef enum CUflushGPUDirectRDMAWritesScope_enum CUflushGPUDirectRDMAWritesScope;
+enum CUflushGPUDirectRDMAWritesTarget_enum : unsigned int {};
+typedef enum CUflushGPUDirectRDMAWritesTarget_enum CUflushGPUDirectRDMAWritesTarget;
+enum CUgraphDebugDot_flags_enum : unsigned int {};
+typedef enum CUgraphDebugDot_flags_enum CUgraphDebugDot_flags;
+enum CUuserObject_flags_enum : unsigned int {};
+typedef enum CUuserObject_flags_enum CUuserObject_flags;
+enum CUuserObjectRetain_flags_enum : unsigned int {};
+typedef enum CUuserObjectRetain_flags_enum CUuserObjectRetain_flags;
+enum CUgraphInstantiate_flags_enum : unsigned int {};
+typedef enum CUgraphInstantiate_flags_enum CUgraphInstantiate_flags;
+enum CUdeviceNumaConfig_enum : unsigned int {};
+typedef enum CUdeviceNumaConfig_enum CUdeviceNumaConfig;
+enum CUmoduleLoadingMode_enum : unsigned int {};
+typedef enum CUmoduleLoadingMode_enum CUmoduleLoadingMode;
+enum CUfunctionLoadingState_enum : unsigned int {};
+typedef enum CUfunctionLoadingState_enum CUfunctionLoadingState;
+enum CUcoredumpSettings_enum : unsigned int {};
+typedef enum CUcoredumpSettings_enum CUcoredumpSettings;
+struct CUgreenCtx_st;
+typedef struct CUgreenCtx_st* CUgreenCtx;
+struct CUdevResourceDesc_st;
+typedef struct CUdevResourceDesc_st* CUdevResourceDesc;
+typedef enum CUgreenCtxCreate_flags CUgreenCtxCreate_flags;
+typedef enum CUdevResourceType CUdevResourceType;
+struct CUdevSmResource_st {
+  unsigned int _0;
+};
+typedef struct CUdevSmResource_st CUdevSmResource;
+struct CUdevResource_st;
+typedef struct CUdevResource_st CUdevResource_v1;
+typedef CUdevResource_v1 CUdevResource;
+
+// Function forward declarations
+CUresult cuGetErrorString(CUresult, const char**);
+CUresult cuGetErrorName(CUresult, const char**);
+CUresult cuInit(unsigned int);
+CUresult cuDriverGetVersion(int*);
+CUresult cuDeviceGet(CUdevice*, int);
+CUresult cuDeviceGetCount(int*);
+CUresult cuDeviceGetName(char*, int, CUdevice);
+CUresult cuDeviceGetUuid(CUuuid*, CUdevice);
+CUresult cuDeviceGetUuid_v2(CUuuid*, CUdevice);
+CUresult cuDeviceGetLuid(char*, unsigned int*, CUdevice);
+CUresult cuDeviceTotalMem_v2(size_t*, CUdevice);
+CUresult cuDeviceGetTexture1DLinearMaxWidth(size_t*, CUarray_format, unsigned int, CUdevice);
+CUresult cuDeviceGetAttribute(int*, CUdevice_attribute, CUdevice);
+CUresult cuDeviceGetNvSciSyncAttributes(void*, CUdevice, int);
+CUresult cuDeviceSetMemPool(CUdevice, CUmemoryPool);
+CUresult cuDeviceGetMemPool(CUmemoryPool*, CUdevice);
+CUresult cuDeviceGetDefaultMemPool(CUmemoryPool*, CUdevice);
+CUresult cuDeviceGetExecAffinitySupport(int*, CUexecAffinityType, CUdevice);
+CUresult cuFlushGPUDirectRDMAWrites(CUflushGPUDirectRDMAWritesTarget, CUflushGPUDirectRDMAWritesScope);
+CUresult cuDeviceGetProperties(CUdevprop*, CUdevice);
+CUresult cuDeviceComputeCapability(int*, int*, CUdevice);
+CUresult cuDevicePrimaryCtxRetain(CUcontext*, CUdevice);
+CUresult cuDevicePrimaryCtxRelease_v2(CUdevice);
+CUresult cuDevicePrimaryCtxSetFlags_v2(CUdevice, unsigned int);
+CUresult cuDevicePrimaryCtxGetState(CUdevice, unsigned int*, int*);
+CUresult cuDevicePrimaryCtxReset_v2(CUdevice);
+CUresult cuCtxCreate_v2(CUcontext*, unsigned int, CUdevice);
+CUresult cuCtxCreate_v3(CUcontext*, CUexecAffinityParam*, int, unsigned int, CUdevice);
+CUresult cuCtxDestroy_v2(CUcontext);
+CUresult cuCtxPushCurrent_v2(CUcontext);
+CUresult cuCtxPopCurrent_v2(CUcontext*);
+CUresult cuCtxSetCurrent(CUcontext);
+CUresult cuCtxGetCurrent(CUcontext*);
+CUresult cuCtxGetDevice(CUdevice*);
+CUresult cuCtxGetFlags(unsigned int*);
+CUresult cuCtxSetFlags(unsigned int);
+CUresult cuCtxGetId(CUcontext, unsigned long long*);
+CUresult cuCtxSynchronize();
+CUresult cuCtxSetLimit(CUlimit, size_t);
+CUresult cuCtxGetLimit(size_t*, CUlimit);
+CUresult cuCtxGetCacheConfig(CUfunc_cache*);
+CUresult cuCtxSetCacheConfig(CUfunc_cache);
+CUresult cuCtxGetApiVersion(CUcontext, unsigned int*);
+CUresult cuCtxGetStreamPriorityRange(int*, int*);
+CUresult cuCtxResetPersistingL2Cache();
+CUresult cuCtxGetExecAffinity(CUexecAffinityParam*, CUexecAffinityType);
+CUresult cuCtxAttach(CUcontext*, unsigned int);
+CUresult cuCtxDetach(CUcontext);
+CUresult cuCtxGetSharedMemConfig(CUsharedconfig*);
+CUresult cuCtxSetSharedMemConfig(CUsharedconfig);
+CUresult cuModuleLoad(CUmodule*, const char*);
+CUresult cuModuleLoadData(CUmodule*, const void*);
+CUresult cuModuleLoadDataEx(CUmodule*, const void*, unsigned int, CUjit_option*, void**);
+CUresult cuModuleLoadFatBinary(CUmodule*, const void*);
+CUresult cuModuleUnload(CUmodule);
+CUresult cuModuleGetLoadingMode(CUmoduleLoadingMode*);
+CUresult cuModuleGetFunction(CUfunction*, CUmodule, const char*);
+CUresult cuModuleGetFunctionCount(unsigned int*, CUmodule);
+CUresult cuModuleEnumerateFunctions(CUfunction*, unsigned int, CUmodule);
+CUresult cuModuleGetGlobal_v2(CUdeviceptr*, size_t*, CUmodule, const char*);
+CUresult cuLinkCreate_v2(unsigned int, CUjit_option*, void**, CUlinkState*);
+CUresult cuLinkAddData_v2(CUlinkState, CUjitInputType, void*, size_t, const char*, unsigned int, CUjit_option*, void**);
+CUresult cuLinkAddFile_v2(CUlinkState, CUjitInputType, const char*, unsigned int, CUjit_option*, void**);
+CUresult cuLinkComplete(CUlinkState, void**, size_t*);
+CUresult cuLinkDestroy(CUlinkState);
+CUresult cuModuleGetTexRef(CUtexref*, CUmodule, const char*);
+CUresult cuModuleGetSurfRef(CUsurfref*, CUmodule, const char*);
+CUresult cuLibraryLoadData(CUlibrary*, const void*, CUjit_option*, void**, unsigned int, CUlibraryOption*, void**, unsigned int);
+CUresult cuLibraryLoadFromFile(CUlibrary*, const char*, CUjit_option*, void**, unsigned int, CUlibraryOption*, void**, unsigned int);
+CUresult cuLibraryUnload(CUlibrary);
+CUresult cuLibraryGetKernel(CUkernel*, CUlibrary, const char*);
+CUresult cuLibraryGetKernelCount(unsigned int*, CUlibrary);
+CUresult cuLibraryEnumerateKernels(CUkernel*, unsigned int, CUlibrary);
+CUresult cuLibraryGetModule(CUmodule*, CUlibrary);
+CUresult cuKernelGetFunction(CUfunction*, CUkernel);
+CUresult cuLibraryGetGlobal(CUdeviceptr*, size_t*, CUlibrary, const char*);
+CUresult cuLibraryGetManaged(CUdeviceptr*, size_t*, CUlibrary, const char*);
+CUresult cuLibraryGetUnifiedFunction(void**, CUlibrary, const char*);
+CUresult cuKernelGetAttribute(int*, CUfunction_attribute, CUkernel, CUdevice);
+CUresult cuKernelSetAttribute(CUfunction_attribute, int, CUkernel, CUdevice);
+CUresult cuKernelSetCacheConfig(CUkernel, CUfunc_cache, CUdevice);
+CUresult cuKernelGetName(const char**, CUkernel);
+CUresult cuKernelGetParamInfo(CUkernel, size_t, size_t*, size_t*);
+CUresult cuMemGetInfo_v2(size_t*, size_t*);
+CUresult cuMemAlloc_v2(CUdeviceptr*, size_t);
+CUresult cuMemAllocPitch_v2(CUdeviceptr*, size_t*, size_t, size_t, unsigned int);
+CUresult cuMemFree_v2(CUdeviceptr);
+CUresult cuMemGetAddressRange_v2(CUdeviceptr*, size_t*, CUdeviceptr);
+CUresult cuMemAllocHost_v2(void**, size_t);
+CUresult cuMemFreeHost(void*);
+CUresult cuMemHostAlloc(void**, size_t, unsigned int);
+CUresult cuMemHostGetDevicePointer_v2(CUdeviceptr*, void*, unsigned int);
+CUresult cuMemHostGetFlags(unsigned int*, void*);
+CUresult cuMemAllocManaged(CUdeviceptr*, size_t, unsigned int);
+// CUresult cuDeviceRegisterAsyncNotification(CUdevice, CUasyncCallback, void *, CUasyncCallbackHandle *);
+CUresult cuDeviceUnregisterAsyncNotification(CUdevice, CUasyncCallbackHandle);
+CUresult cuDeviceGetByPCIBusId(CUdevice*, const char*);
+CUresult cuDeviceGetPCIBusId(char*, int, CUdevice);
+CUresult cuIpcGetEventHandle(CUipcEventHandle*, CUevent);
+CUresult cuIpcOpenEventHandle(CUevent*, CUipcEventHandle);
+CUresult cuIpcGetMemHandle(CUipcMemHandle*, CUdeviceptr);
+CUresult cuIpcOpenMemHandle_v2(CUdeviceptr*, CUipcMemHandle, unsigned int);
+CUresult cuIpcCloseMemHandle(CUdeviceptr);
+CUresult cuMemHostRegister_v2(void*, size_t, unsigned int);
+CUresult cuMemHostUnregister(void*);
+CUresult cuMemcpy(CUdeviceptr, CUdeviceptr, size_t);
+CUresult cuMemcpyPeer(CUdeviceptr, CUcontext, CUdeviceptr, CUcontext, size_t);
+CUresult cuMemcpyHtoD_v2(CUdeviceptr, const void*, size_t);
+CUresult cuMemcpyDtoH_v2(void*, CUdeviceptr, size_t);
+CUresult cuMemcpyDtoD_v2(CUdeviceptr, CUdeviceptr, size_t);
+CUresult cuMemcpyDtoA_v2(CUarray, size_t, CUdeviceptr, size_t);
+CUresult cuMemcpyAtoD_v2(CUdeviceptr, CUarray, size_t, size_t);
+CUresult cuMemcpyHtoA_v2(CUarray, size_t, const void*, size_t);
+CUresult cuMemcpyAtoH_v2(void*, CUarray, size_t, size_t);
+CUresult cuMemcpyAtoA_v2(CUarray, size_t, CUarray, size_t, size_t);
+CUresult cuMemcpy2D_v2(const CUDA_MEMCPY2D*);
+CUresult cuMemcpy2DUnaligned_v2(const CUDA_MEMCPY2D*);
+CUresult cuMemcpy3D_v2(const CUDA_MEMCPY3D*);
+CUresult cuMemcpy3DPeer(const CUDA_MEMCPY3D_PEER*);
+CUresult cuMemcpyAsync(CUdeviceptr, CUdeviceptr, size_t, CUstream);
+CUresult cuMemcpyPeerAsync(CUdeviceptr, CUcontext, CUdeviceptr, CUcontext, size_t, CUstream);
+CUresult cuMemcpyHtoDAsync_v2(CUdeviceptr, const void*, size_t, CUstream);
+CUresult cuMemcpyDtoHAsync_v2(void*, CUdeviceptr, size_t, CUstream);
+CUresult cuMemcpyDtoDAsync_v2(CUdeviceptr, CUdeviceptr, size_t, CUstream);
+CUresult cuMemcpyHtoAAsync_v2(CUarray, size_t, const void*, size_t, CUstream);
+CUresult cuMemcpyAtoHAsync_v2(void*, CUarray, size_t, size_t, CUstream);
+CUresult cuMemcpy2DAsync_v2(const CUDA_MEMCPY2D*, CUstream);
+CUresult cuMemcpy3DAsync_v2(const CUDA_MEMCPY3D*, CUstream);
+CUresult cuMemcpy3DPeerAsync(const CUDA_MEMCPY3D_PEER*, CUstream);
+CUresult cuMemsetD8_v2(CUdeviceptr, unsigned char, size_t);
+CUresult cuMemsetD16_v2(CUdeviceptr, unsigned short, size_t);
+CUresult cuMemsetD32_v2(CUdeviceptr, unsigned int, size_t);
+CUresult cuMemsetD2D8_v2(CUdeviceptr, size_t, unsigned char, size_t, size_t);
+CUresult cuMemsetD2D16_v2(CUdeviceptr, size_t, unsigned short, size_t, size_t);
+CUresult cuMemsetD2D32_v2(CUdeviceptr, size_t, unsigned int, size_t, size_t);
+CUresult cuMemsetD8Async(CUdeviceptr, unsigned char, size_t, CUstream);
+CUresult cuMemsetD16Async(CUdeviceptr, unsigned short, size_t, CUstream);
+CUresult cuMemsetD32Async(CUdeviceptr, unsigned int, size_t, CUstream);
+CUresult cuMemsetD2D8Async(CUdeviceptr, size_t, unsigned char, size_t, size_t, CUstream);
+CUresult cuMemsetD2D16Async(CUdeviceptr, size_t, unsigned short, size_t, size_t, CUstream);
+CUresult cuMemsetD2D32Async(CUdeviceptr, size_t, unsigned int, size_t, size_t, CUstream);
+CUresult cuArrayCreate_v2(CUarray*, const CUDA_ARRAY_DESCRIPTOR*);
+CUresult cuArrayGetDescriptor_v2(CUDA_ARRAY_DESCRIPTOR*, CUarray);
+CUresult cuArrayGetSparseProperties(CUDA_ARRAY_SPARSE_PROPERTIES*, CUarray);
+CUresult cuMipmappedArrayGetSparseProperties(CUDA_ARRAY_SPARSE_PROPERTIES*, CUmipmappedArray);
+CUresult cuArrayGetMemoryRequirements(CUDA_ARRAY_MEMORY_REQUIREMENTS*, CUarray, CUdevice);
+CUresult cuMipmappedArrayGetMemoryRequirements(CUDA_ARRAY_MEMORY_REQUIREMENTS*, CUmipmappedArray, CUdevice);
+CUresult cuArrayGetPlane(CUarray*, CUarray, unsigned int);
+CUresult cuArrayDestroy(CUarray);
+CUresult cuArray3DCreate_v2(CUarray*, const CUDA_ARRAY3D_DESCRIPTOR*);
+CUresult cuArray3DGetDescriptor_v2(CUDA_ARRAY3D_DESCRIPTOR*, CUarray);
+CUresult cuMipmappedArrayCreate(CUmipmappedArray*, const CUDA_ARRAY3D_DESCRIPTOR*, unsigned int);
+CUresult cuMipmappedArrayGetLevel(CUarray*, CUmipmappedArray, unsigned int);
+CUresult cuMipmappedArrayDestroy(CUmipmappedArray);
+CUresult cuMemGetHandleForAddressRange(void*, CUdeviceptr, size_t, CUmemRangeHandleType, unsigned long long);
+CUresult cuMemAddressReserve(CUdeviceptr*, size_t, size_t, CUdeviceptr, unsigned long long);
+CUresult cuMemAddressFree(CUdeviceptr, size_t);
+CUresult cuMemCreate(CUmemGenericAllocationHandle*, size_t, const CUmemAllocationProp*, unsigned long long);
+CUresult cuMemRelease(CUmemGenericAllocationHandle);
+CUresult cuMemMap(CUdeviceptr, size_t, size_t, CUmemGenericAllocationHandle, unsigned long long);
+CUresult cuMemMapArrayAsync(CUarrayMapInfo*, unsigned int, CUstream);
+CUresult cuMemUnmap(CUdeviceptr, size_t);
+CUresult cuMemSetAccess(CUdeviceptr, size_t, const CUmemAccessDesc*, size_t);
+CUresult cuMemGetAccess(unsigned long long*, const CUmemLocation*, CUdeviceptr);
+CUresult cuMemExportToShareableHandle(void*, CUmemGenericAllocationHandle, CUmemAllocationHandleType, unsigned long long);
+CUresult cuMemImportFromShareableHandle(CUmemGenericAllocationHandle*, void*, CUmemAllocationHandleType);
+CUresult cuMemGetAllocationGranularity(size_t*, const CUmemAllocationProp*, CUmemAllocationGranularity_flags);
+CUresult cuMemGetAllocationPropertiesFromHandle(CUmemAllocationProp*, CUmemGenericAllocationHandle);
+CUresult cuMemRetainAllocationHandle(CUmemGenericAllocationHandle*, void*);
+CUresult cuMemFreeAsync(CUdeviceptr, CUstream);
+CUresult cuMemAllocAsync(CUdeviceptr*, size_t, CUstream);
+CUresult cuMemPoolTrimTo(CUmemoryPool, size_t);
+CUresult cuMemPoolSetAttribute(CUmemoryPool, CUmemPool_attribute, void*);
+CUresult cuMemPoolGetAttribute(CUmemoryPool, CUmemPool_attribute, void*);
+CUresult cuMemPoolSetAccess(CUmemoryPool, const CUmemAccessDesc*, size_t);
+CUresult cuMemPoolGetAccess(CUmemAccess_flags*, CUmemoryPool, CUmemLocation*);
+CUresult cuMemPoolCreate(CUmemoryPool*, const CUmemPoolProps*);
+CUresult cuMemPoolDestroy(CUmemoryPool);
+CUresult cuMemAllocFromPoolAsync(CUdeviceptr*, size_t, CUmemoryPool, CUstream);
+CUresult cuMemPoolExportToShareableHandle(void*, CUmemoryPool, CUmemAllocationHandleType, unsigned long long);
+CUresult cuMemPoolImportFromShareableHandle(CUmemoryPool*, void*, CUmemAllocationHandleType, unsigned long long);
+CUresult cuMemPoolExportPointer(CUmemPoolPtrExportData*, CUdeviceptr);
+CUresult cuMemPoolImportPointer(CUdeviceptr*, CUmemoryPool, CUmemPoolPtrExportData*);
+CUresult cuMulticastCreate(CUmemGenericAllocationHandle*, const CUmulticastObjectProp*);
+CUresult cuMulticastAddDevice(CUmemGenericAllocationHandle, CUdevice);
+CUresult cuMulticastBindMem(CUmemGenericAllocationHandle, size_t, CUmemGenericAllocationHandle, size_t, size_t, unsigned long long);
+CUresult cuMulticastBindAddr(CUmemGenericAllocationHandle, size_t, CUdeviceptr, size_t, unsigned long long);
+CUresult cuMulticastUnbind(CUmemGenericAllocationHandle, CUdevice, size_t, size_t);
+CUresult cuMulticastGetGranularity(size_t*, const CUmulticastObjectProp*, CUmulticastGranularity_flags);
+CUresult cuPointerGetAttribute(void*, CUpointer_attribute, CUdeviceptr);
+CUresult cuMemPrefetchAsync(CUdeviceptr, size_t, CUdevice, CUstream);
+CUresult cuMemPrefetchAsync_v2(CUdeviceptr, size_t, CUmemLocation, unsigned int, CUstream);
+CUresult cuMemAdvise(CUdeviceptr, size_t, CUmem_advise, CUdevice);
+CUresult cuMemAdvise_v2(CUdeviceptr, size_t, CUmem_advise, CUmemLocation);
+CUresult cuMemRangeGetAttribute(void*, size_t, CUmem_range_attribute, CUdeviceptr, size_t);
+CUresult cuMemRangeGetAttributes(void**, size_t*, CUmem_range_attribute*, size_t, CUdeviceptr, size_t);
+CUresult cuPointerSetAttribute(const void*, CUpointer_attribute, CUdeviceptr);
+CUresult cuPointerGetAttributes(unsigned int, CUpointer_attribute*, void**, CUdeviceptr);
+CUresult cuStreamCreate(CUstream*, unsigned int);
+CUresult cuStreamCreateWithPriority(CUstream*, unsigned int, int);
+CUresult cuStreamGetPriority(CUstream, int*);
+CUresult cuStreamGetFlags(CUstream, unsigned int*);
+CUresult cuStreamGetId(CUstream, unsigned long long*);
+CUresult cuStreamGetCtx(CUstream, CUcontext*);
+CUresult cuStreamWaitEvent(CUstream, CUevent, unsigned int);
+// CUresult cuStreamAddCallback(CUstream, CUstreamCallback, void *, unsigned int);
+CUresult cuStreamBeginCapture_v2(CUstream, CUstreamCaptureMode);
+CUresult cuStreamBeginCaptureToGraph(CUstream, CUgraph, const CUgraphNode*, const CUgraphEdgeData*, size_t, CUstreamCaptureMode);
+CUresult cuThreadExchangeStreamCaptureMode(CUstreamCaptureMode*);
+CUresult cuStreamEndCapture(CUstream, CUgraph*);
+CUresult cuStreamIsCapturing(CUstream, CUstreamCaptureStatus*);
+// CUresult cuStreamGetCaptureInfo_v2(CUstream, CUstreamCaptureStatus *, cuuint64_t *, CUgraph *, const CUgraphNode **, size_t *);
+// CUresult cuStreamGetCaptureInfo_v3(CUstream, CUstreamCaptureStatus *, cuuint64_t *, CUgraph *, const CUgraphNode **, const CUgraphEdgeData **, size_t *);
+CUresult cuStreamUpdateCaptureDependencies(CUstream, CUgraphNode*, size_t, unsigned int);
+CUresult cuStreamUpdateCaptureDependencies_v2(CUstream, CUgraphNode*, const CUgraphEdgeData*, size_t, unsigned int);
+CUresult cuStreamAttachMemAsync(CUstream, CUdeviceptr, size_t, unsigned int);
+CUresult cuStreamQuery(CUstream);
+CUresult cuStreamSynchronize(CUstream);
+CUresult cuStreamDestroy_v2(CUstream);
+CUresult cuStreamCopyAttributes(CUstream, CUstream);
+CUresult cuStreamGetAttribute(CUstream, CUstreamAttrID, CUstreamAttrValue*);
+CUresult cuStreamSetAttribute(CUstream, CUstreamAttrID, const CUstreamAttrValue*);
+CUresult cuEventCreate(CUevent*, unsigned int);
+CUresult cuEventRecord(CUevent, CUstream);
+CUresult cuEventRecordWithFlags(CUevent, CUstream, unsigned int);
+CUresult cuEventQuery(CUevent);
+CUresult cuEventSynchronize(CUevent);
+CUresult cuEventDestroy_v2(CUevent);
+CUresult cuEventElapsedTime(float*, CUevent, CUevent);
+CUresult cuImportExternalMemory(CUexternalMemory*, const CUDA_EXTERNAL_MEMORY_HANDLE_DESC*);
+CUresult cuExternalMemoryGetMappedBuffer(CUdeviceptr*, CUexternalMemory, const CUDA_EXTERNAL_MEMORY_BUFFER_DESC*);
+CUresult cuExternalMemoryGetMappedMipmappedArray(CUmipmappedArray*, CUexternalMemory, const CUDA_EXTERNAL_MEMORY_MIPMAPPED_ARRAY_DESC*);
+CUresult cuDestroyExternalMemory(CUexternalMemory);
+CUresult cuImportExternalSemaphore(CUexternalSemaphore*, const CUDA_EXTERNAL_SEMAPHORE_HANDLE_DESC*);
+CUresult cuSignalExternalSemaphoresAsync(const CUexternalSemaphore*, const CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS*, unsigned int, CUstream);
+CUresult cuWaitExternalSemaphoresAsync(const CUexternalSemaphore*, const CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS*, unsigned int, CUstream);
+CUresult cuDestroyExternalSemaphore(CUexternalSemaphore);
+CUresult cuStreamWaitValue32_v2(CUstream, CUdeviceptr, cuuint32_t, unsigned int);
+CUresult cuStreamWaitValue64_v2(CUstream, CUdeviceptr, cuuint64_t, unsigned int);
+CUresult cuStreamWriteValue32_v2(CUstream, CUdeviceptr, cuuint32_t, unsigned int);
+CUresult cuStreamWriteValue64_v2(CUstream, CUdeviceptr, cuuint64_t, unsigned int);
+CUresult cuStreamBatchMemOp_v2(CUstream, unsigned int, CUstreamBatchMemOpParams*, unsigned int);
+CUresult cuFuncGetAttribute(int*, CUfunction_attribute, CUfunction);
+CUresult cuFuncSetAttribute(CUfunction, CUfunction_attribute, int);
+CUresult cuFuncSetCacheConfig(CUfunction, CUfunc_cache);
+CUresult cuFuncGetModule(CUmodule*, CUfunction);
+CUresult cuFuncGetName(const char**, CUfunction);
+CUresult cuFuncGetParamInfo(CUfunction, size_t, size_t*, size_t*);
+CUresult cuFuncIsLoaded(CUfunctionLoadingState*, CUfunction);
+CUresult cuFuncLoad(CUfunction);
+CUresult cuLaunchKernel(CUfunction, unsigned int, unsigned int, unsigned int, unsigned int, unsigned int, unsigned int, unsigned int,
+                        CUstream, void**, void**);
+CUresult cuLaunchKernelEx(const CUlaunchConfig*, CUfunction, void**, void**);
+CUresult cuLaunchCooperativeKernel(CUfunction, unsigned int, unsigned int, unsigned int, unsigned int, unsigned int, unsigned int,
+                                   unsigned int, CUstream, void**);
+CUresult cuLaunchCooperativeKernelMultiDevice(CUDA_LAUNCH_PARAMS*, unsigned int, unsigned int);
+CUresult cuLaunchHostFunc(CUstream, CUhostFn, void*);
+CUresult cuFuncSetBlockShape(CUfunction, int, int, int);
+CUresult cuFuncSetSharedSize(CUfunction, unsigned int);
+CUresult cuParamSetSize(CUfunction, unsigned int);
+CUresult cuParamSeti(CUfunction, int, unsigned int);
+CUresult cuParamSetf(CUfunction, int, float);
+CUresult cuParamSetv(CUfunction, int, void*, unsigned int);
+CUresult cuLaunch(CUfunction);
+CUresult cuLaunchGrid(CUfunction, int, int);
+CUresult cuLaunchGridAsync(CUfunction, int, int, CUstream);
+CUresult cuParamSetTexRef(CUfunction, int, CUtexref);
+CUresult cuFuncSetSharedMemConfig(CUfunction, CUsharedconfig);
+CUresult cuGraphCreate(CUgraph*, unsigned int);
+CUresult cuGraphAddKernelNode_v2(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, const CUDA_KERNEL_NODE_PARAMS*);
+CUresult cuGraphKernelNodeGetParams_v2(CUgraphNode, CUDA_KERNEL_NODE_PARAMS*);
+CUresult cuGraphKernelNodeSetParams_v2(CUgraphNode, const CUDA_KERNEL_NODE_PARAMS*);
+CUresult cuGraphAddMemcpyNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, const CUDA_MEMCPY3D*, CUcontext);
+CUresult cuGraphMemcpyNodeGetParams(CUgraphNode, CUDA_MEMCPY3D*);
+CUresult cuGraphMemcpyNodeSetParams(CUgraphNode, const CUDA_MEMCPY3D*);
+CUresult cuGraphAddMemsetNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, const CUDA_MEMSET_NODE_PARAMS*, CUcontext);
+CUresult cuGraphMemsetNodeGetParams(CUgraphNode, CUDA_MEMSET_NODE_PARAMS*);
+CUresult cuGraphMemsetNodeSetParams(CUgraphNode, const CUDA_MEMSET_NODE_PARAMS*);
+CUresult cuGraphAddHostNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, const CUDA_HOST_NODE_PARAMS*);
+CUresult cuGraphHostNodeGetParams(CUgraphNode, CUDA_HOST_NODE_PARAMS*);
+CUresult cuGraphHostNodeSetParams(CUgraphNode, const CUDA_HOST_NODE_PARAMS*);
+CUresult cuGraphAddChildGraphNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, CUgraph);
+CUresult cuGraphChildGraphNodeGetGraph(CUgraphNode, CUgraph*);
+CUresult cuGraphAddEmptyNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t);
+CUresult cuGraphAddEventRecordNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, CUevent);
+CUresult cuGraphEventRecordNodeGetEvent(CUgraphNode, CUevent*);
+CUresult cuGraphEventRecordNodeSetEvent(CUgraphNode, CUevent);
+CUresult cuGraphAddEventWaitNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, CUevent);
+CUresult cuGraphEventWaitNodeGetEvent(CUgraphNode, CUevent*);
+CUresult cuGraphEventWaitNodeSetEvent(CUgraphNode, CUevent);
+CUresult cuGraphAddExternalSemaphoresSignalNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, const CUDA_EXT_SEM_SIGNAL_NODE_PARAMS*);
+CUresult cuGraphExternalSemaphoresSignalNodeGetParams(CUgraphNode, CUDA_EXT_SEM_SIGNAL_NODE_PARAMS*);
+CUresult cuGraphExternalSemaphoresSignalNodeSetParams(CUgraphNode, const CUDA_EXT_SEM_SIGNAL_NODE_PARAMS*);
+CUresult cuGraphAddExternalSemaphoresWaitNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, const CUDA_EXT_SEM_WAIT_NODE_PARAMS*);
+CUresult cuGraphExternalSemaphoresWaitNodeGetParams(CUgraphNode, CUDA_EXT_SEM_WAIT_NODE_PARAMS*);
+CUresult cuGraphExternalSemaphoresWaitNodeSetParams(CUgraphNode, const CUDA_EXT_SEM_WAIT_NODE_PARAMS*);
+CUresult cuGraphAddBatchMemOpNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, const CUDA_BATCH_MEM_OP_NODE_PARAMS*);
+CUresult cuGraphBatchMemOpNodeGetParams(CUgraphNode, CUDA_BATCH_MEM_OP_NODE_PARAMS*);
+CUresult cuGraphBatchMemOpNodeSetParams(CUgraphNode, const CUDA_BATCH_MEM_OP_NODE_PARAMS*);
+CUresult cuGraphExecBatchMemOpNodeSetParams(CUgraphExec, CUgraphNode, const CUDA_BATCH_MEM_OP_NODE_PARAMS*);
+CUresult cuGraphAddMemAllocNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, CUDA_MEM_ALLOC_NODE_PARAMS*);
+CUresult cuGraphMemAllocNodeGetParams(CUgraphNode, CUDA_MEM_ALLOC_NODE_PARAMS*);
+CUresult cuGraphAddMemFreeNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, CUdeviceptr);
+CUresult cuGraphMemFreeNodeGetParams(CUgraphNode, CUdeviceptr*);
+CUresult cuDeviceGraphMemTrim(CUdevice);
+CUresult cuDeviceGetGraphMemAttribute(CUdevice, CUgraphMem_attribute, void*);
+CUresult cuDeviceSetGraphMemAttribute(CUdevice, CUgraphMem_attribute, void*);
+CUresult cuGraphClone(CUgraph*, CUgraph);
+CUresult cuGraphNodeFindInClone(CUgraphNode*, CUgraphNode, CUgraph);
+CUresult cuGraphNodeGetType(CUgraphNode, CUgraphNodeType*);
+CUresult cuGraphGetNodes(CUgraph, CUgraphNode*, size_t*);
+CUresult cuGraphGetRootNodes(CUgraph, CUgraphNode*, size_t*);
+CUresult cuGraphGetEdges(CUgraph, CUgraphNode*, CUgraphNode*, size_t*);
+CUresult cuGraphGetEdges_v2(CUgraph, CUgraphNode*, CUgraphNode*, CUgraphEdgeData*, size_t*);
+CUresult cuGraphNodeGetDependencies(CUgraphNode, CUgraphNode*, size_t*);
+CUresult cuGraphNodeGetDependencies_v2(CUgraphNode, CUgraphNode*, CUgraphEdgeData*, size_t*);
+CUresult cuGraphNodeGetDependentNodes(CUgraphNode, CUgraphNode*, size_t*);
+CUresult cuGraphNodeGetDependentNodes_v2(CUgraphNode, CUgraphNode*, CUgraphEdgeData*, size_t*);
+CUresult cuGraphAddDependencies(CUgraph, const CUgraphNode*, const CUgraphNode*, size_t);
+CUresult cuGraphAddDependencies_v2(CUgraph, const CUgraphNode*, const CUgraphNode*, const CUgraphEdgeData*, size_t);
+CUresult cuGraphRemoveDependencies(CUgraph, const CUgraphNode*, const CUgraphNode*, size_t);
+CUresult cuGraphRemoveDependencies_v2(CUgraph, const CUgraphNode*, const CUgraphNode*, const CUgraphEdgeData*, size_t);
+CUresult cuGraphDestroyNode(CUgraphNode);
+CUresult cuGraphInstantiateWithFlags(CUgraphExec*, CUgraph, unsigned long long);
+CUresult cuGraphInstantiateWithParams(CUgraphExec*, CUgraph, CUDA_GRAPH_INSTANTIATE_PARAMS*);
+CUresult cuGraphExecGetFlags(CUgraphExec, cuuint64_t*);
+CUresult cuGraphExecKernelNodeSetParams_v2(CUgraphExec, CUgraphNode, const CUDA_KERNEL_NODE_PARAMS*);
+CUresult cuGraphExecMemcpyNodeSetParams(CUgraphExec, CUgraphNode, const CUDA_MEMCPY3D*, CUcontext);
+CUresult cuGraphExecMemsetNodeSetParams(CUgraphExec, CUgraphNode, const CUDA_MEMSET_NODE_PARAMS*, CUcontext);
+CUresult cuGraphExecHostNodeSetParams(CUgraphExec, CUgraphNode, const CUDA_HOST_NODE_PARAMS*);
+CUresult cuGraphExecChildGraphNodeSetParams(CUgraphExec, CUgraphNode, CUgraph);
+CUresult cuGraphExecEventRecordNodeSetEvent(CUgraphExec, CUgraphNode, CUevent);
+CUresult cuGraphExecEventWaitNodeSetEvent(CUgraphExec, CUgraphNode, CUevent);
+CUresult cuGraphExecExternalSemaphoresSignalNodeSetParams(CUgraphExec, CUgraphNode, const CUDA_EXT_SEM_SIGNAL_NODE_PARAMS*);
+CUresult cuGraphExecExternalSemaphoresWaitNodeSetParams(CUgraphExec, CUgraphNode, const CUDA_EXT_SEM_WAIT_NODE_PARAMS*);
+CUresult cuGraphNodeSetEnabled(CUgraphExec, CUgraphNode, unsigned int);
+CUresult cuGraphNodeGetEnabled(CUgraphExec, CUgraphNode, unsigned int*);
+CUresult cuGraphUpload(CUgraphExec, CUstream);
+CUresult cuGraphLaunch(CUgraphExec, CUstream);
+CUresult cuGraphExecDestroy(CUgraphExec);
+CUresult cuGraphDestroy(CUgraph);
+CUresult cuGraphExecUpdate_v2(CUgraphExec, CUgraph, CUgraphExecUpdateResultInfo*);
+CUresult cuGraphKernelNodeCopyAttributes(CUgraphNode, CUgraphNode);
+CUresult cuGraphKernelNodeGetAttribute(CUgraphNode, CUkernelNodeAttrID, CUkernelNodeAttrValue*);
+CUresult cuGraphKernelNodeSetAttribute(CUgraphNode, CUkernelNodeAttrID, const CUkernelNodeAttrValue*);
+CUresult cuGraphDebugDotPrint(CUgraph, const char*, unsigned int);
+CUresult cuUserObjectCreate(CUuserObject*, void*, CUhostFn, unsigned int, unsigned int);
+CUresult cuUserObjectRetain(CUuserObject, unsigned int);
+CUresult cuUserObjectRelease(CUuserObject, unsigned int);
+CUresult cuGraphRetainUserObject(CUgraph, CUuserObject, unsigned int, unsigned int);
+CUresult cuGraphReleaseUserObject(CUgraph, CUuserObject, unsigned int);
+CUresult cuGraphAddNode(CUgraphNode*, CUgraph, const CUgraphNode*, size_t, CUgraphNodeParams*);
+CUresult cuGraphAddNode_v2(CUgraphNode*, CUgraph, const CUgraphNode*, const CUgraphEdgeData*, size_t, CUgraphNodeParams*);
+CUresult cuGraphNodeSetParams(CUgraphNode, CUgraphNodeParams*);
+CUresult cuGraphExecNodeSetParams(CUgraphExec, CUgraphNode, CUgraphNodeParams*);
+CUresult cuGraphConditionalHandleCreate(CUgraphConditionalHandle*, CUgraph, CUcontext, unsigned int, unsigned int);
+CUresult cuOccupancyMaxActiveBlocksPerMultiprocessor(int*, CUfunction, int, size_t);
+CUresult cuOccupancyMaxActiveBlocksPerMultiprocessorWithFlags(int*, CUfunction, int, size_t, unsigned int);
+// CUresult cuOccupancyMaxPotentialBlockSize(int *, int *, CUfunction, CUoccupancyB2DSize, size_t, int);
+// CUresult cuOccupancyMaxPotentialBlockSizeWithFlags(int *, int *, CUfunction, CUoccupancyB2DSize, size_t, int, unsigned int);
+CUresult cuOccupancyAvailableDynamicSMemPerBlock(size_t*, CUfunction, int, int);
+CUresult cuOccupancyMaxPotentialClusterSize(int*, CUfunction, const CUlaunchConfig*);
+CUresult cuOccupancyMaxActiveClusters(int*, CUfunction, const CUlaunchConfig*);
+CUresult cuTexRefSetArray(CUtexref, CUarray, unsigned int);
+CUresult cuTexRefSetMipmappedArray(CUtexref, CUmipmappedArray, unsigned int);
+CUresult cuTexRefSetAddress_v2(size_t*, CUtexref, CUdeviceptr, size_t);
+CUresult cuTexRefSetAddress2D_v3(CUtexref, const CUDA_ARRAY_DESCRIPTOR*, CUdeviceptr, size_t);
+CUresult cuTexRefSetFormat(CUtexref, CUarray_format, int);
+CUresult cuTexRefSetAddressMode(CUtexref, int, CUaddress_mode);
+CUresult cuTexRefSetFilterMode(CUtexref, CUfilter_mode);
+CUresult cuTexRefSetMipmapFilterMode(CUtexref, CUfilter_mode);
+CUresult cuTexRefSetMipmapLevelBias(CUtexref, float);
+CUresult cuTexRefSetMipmapLevelClamp(CUtexref, float, float);
+CUresult cuTexRefSetMaxAnisotropy(CUtexref, unsigned int);
+CUresult cuTexRefSetBorderColor(CUtexref, float*);
+CUresult cuTexRefSetFlags(CUtexref, unsigned int);
+CUresult cuTexRefGetAddress_v2(CUdeviceptr*, CUtexref);
+CUresult cuTexRefGetArray(CUarray*, CUtexref);
+CUresult cuTexRefGetMipmappedArray(CUmipmappedArray*, CUtexref);
+CUresult cuTexRefGetAddressMode(CUaddress_mode*, CUtexref, int);
+CUresult cuTexRefGetFilterMode(CUfilter_mode*, CUtexref);
+CUresult cuTexRefGetFormat(CUarray_format*, int*, CUtexref);
+CUresult cuTexRefGetMipmapFilterMode(CUfilter_mode*, CUtexref);
+CUresult cuTexRefGetMipmapLevelBias(float*, CUtexref);
+CUresult cuTexRefGetMipmapLevelClamp(float*, float*, CUtexref);
+CUresult cuTexRefGetMaxAnisotropy(int*, CUtexref);
+CUresult cuTexRefGetBorderColor(float*, CUtexref);
+CUresult cuTexRefGetFlags(unsigned int*, CUtexref);
+CUresult cuTexRefCreate(CUtexref*);
+CUresult cuTexRefDestroy(CUtexref);
+CUresult cuSurfRefSetArray(CUsurfref, CUarray, unsigned int);
+CUresult cuSurfRefGetArray(CUarray*, CUsurfref);
+CUresult cuTexObjectCreate(CUtexObject*, const CUDA_RESOURCE_DESC*, const CUDA_TEXTURE_DESC*, const CUDA_RESOURCE_VIEW_DESC*);
+CUresult cuTexObjectDestroy(CUtexObject);
+CUresult cuTexObjectGetResourceDesc(CUDA_RESOURCE_DESC*, CUtexObject);
+CUresult cuTexObjectGetTextureDesc(CUDA_TEXTURE_DESC*, CUtexObject);
+CUresult cuTexObjectGetResourceViewDesc(CUDA_RESOURCE_VIEW_DESC*, CUtexObject);
+CUresult cuSurfObjectCreate(CUsurfObject*, const CUDA_RESOURCE_DESC*);
+CUresult cuSurfObjectDestroy(CUsurfObject);
+CUresult cuSurfObjectGetResourceDesc(CUDA_RESOURCE_DESC*, CUsurfObject);
+CUresult cuTensorMapEncodeTiled(CUtensorMap*, CUtensorMapDataType, cuuint32_t, void*, const cuuint64_t*, const cuuint64_t*, const cuuint32_t*,
+                                const cuuint32_t*, CUtensorMapInterleave, CUtensorMapSwizzle, CUtensorMapL2promotion, CUtensorMapFloatOOBfill);
+// CUresult cuTensorMapEncodeIm2col(CUtensorMap *, CUtensorMapDataType, cuuint32_t, void *, const cuuint64_t *, const cuuint64_t *, const int
+// *, const int *, cuuint32_t, cuuint32_t, const cuuint32_t *, CUtensorMapInterleave, CUtensorMapSwizzle, CUtensorMapL2promotion, CUtensorMapFloatOOBfill);
+CUresult cuTensorMapReplaceAddress(CUtensorMap*, void*);
+CUresult cuDeviceCanAccessPeer(int*, CUdevice, CUdevice);
+CUresult cuCtxEnablePeerAccess(CUcontext, unsigned int);
+CUresult cuCtxDisablePeerAccess(CUcontext);
+CUresult cuDeviceGetP2PAttribute(int*, CUdevice_P2PAttribute, CUdevice, CUdevice);
+CUresult cuGraphicsUnregisterResource(CUgraphicsResource);
+CUresult cuGraphicsSubResourceGetMappedArray(CUarray*, CUgraphicsResource, unsigned int, unsigned int);
+CUresult cuGraphicsResourceGetMappedMipmappedArray(CUmipmappedArray*, CUgraphicsResource);
+CUresult cuGraphicsResourceGetMappedPointer_v2(CUdeviceptr*, size_t*, CUgraphicsResource);
+CUresult cuGraphicsResourceSetMapFlags_v2(CUgraphicsResource, unsigned int);
+CUresult cuGraphicsMapResources(unsigned int, CUgraphicsResource*, CUstream);
+CUresult cuGraphicsUnmapResources(unsigned int, CUgraphicsResource*, CUstream);
+CUresult cuGetProcAddress_v2(const char*, void**, int, cuuint64_t, CUdriverProcAddressQueryResult*);
+CUresult cuCoredumpGetAttribute(CUcoredumpSettings, void*, size_t*);
+CUresult cuCoredumpGetAttributeGlobal(CUcoredumpSettings, void*, size_t*);
+CUresult cuCoredumpSetAttribute(CUcoredumpSettings, void*, size_t*);
+CUresult cuCoredumpSetAttributeGlobal(CUcoredumpSettings, void*, size_t*);
+CUresult cuGetExportTable(const void**, const CUuuid*);
+CUresult cuGreenCtxCreate(CUgreenCtx*, CUdevResourceDesc, CUdevice, unsigned int);
+CUresult cuGreenCtxDestroy(CUgreenCtx);
+CUresult cuCtxFromGreenCtx(CUcontext*, CUgreenCtx);
+CUresult cuDeviceGetDevResource(CUdevice, CUdevResource*, CUdevResourceType);
+CUresult cuCtxGetDevResource(CUcontext, CUdevResource*, CUdevResourceType);
+CUresult cuGreenCtxGetDevResource(CUgreenCtx, CUdevResource*, CUdevResourceType);
+CUresult cuDevSmResourceSplitByCount(CUdevResource*, unsigned int*, const CUdevResource*, CUdevResource*, unsigned int, unsigned int);
+CUresult cuDevResourceGenerateDesc(CUdevResourceDesc*, CUdevResource*, unsigned int);
+CUresult cuGreenCtxRecordEvent(CUgreenCtx, CUevent);
+CUresult cuGreenCtxWaitEvent(CUgreenCtx, CUevent);
+CUresult cuStreamGetGreenCtx(CUstream, CUgreenCtx*);
+#ifdef __cplusplus
+}
+#endif
diff --git a/ThunkLibs/libcuda/fixup_defines.h b/ThunkLibs/libcuda/fixup_defines.h
new file mode 100644
index 000000000..cc8a94990
--- /dev/null
+++ b/ThunkLibs/libcuda/fixup_defines.h
@@ -0,0 +1,15 @@
+// SPDX-License-Identifier: MIT
+#pragma once
+#include <cstddef>
+#ifdef __cplusplus
+extern "C" {
+#endif
+
+typedef void (*CUhostFn)(void*);
+enum cudaError_enum {};
+enum CUgreenCtxCreate_flags {};
+enum CUdevResourceType {};
+
+#ifdef __cplusplus
+}
+#endif
diff --git a/ThunkLibs/libcuda/libcuda_Guest.cpp b/ThunkLibs/libcuda/libcuda_Guest.cpp
new file mode 100644
index 000000000..beac01ed9
--- /dev/null
+++ b/ThunkLibs/libcuda/libcuda_Guest.cpp
@@ -0,0 +1,84 @@
+// SPDX-License-Identifier: MIT
+#include "common/Guest.h"
+#include "cuda_defines.h"
+
+#include "thunkgen_guest_libcuda.inl"
+#include <cstdio>
+#include <dlfcn.h>
+#include <functional>
+#include <string_view>
+#include <unordered_map>
+
+#include <cstdint>
+
+// Maps cuda API function names to the address of a guest function which is
+// linked to the corresponding host function pointer
+const std::unordered_map<std::string_view, uintptr_t /* guest function address */> HostPtrInvokers = std::invoke([]() {
+#define PAIR(name, unused) Ret[#name] = reinterpret_cast<uintptr_t>(GetCallerForHostFunction(name));
+  std::unordered_map<std::string_view, uintptr_t> Ret;
+  FOREACH_internal_SYMBOL(PAIR);
+  return Ret;
+#undef PAIR
+});
+
+extern "C" {
+
+// This variable controls the behavior of cuGetProcAddress for functions we don't know the signature of:
+// - if false (default), we return a nullptr (since the application might have a fallback code path)
+// - if true, we return a stub function that fatally errors upon being called
+constexpr bool stub_unknown_functions = false;
+
+// Fatally erroring function with a thunk-like interface. This is used as a placeholder for unknown CUDA functions
+[[noreturn]]
+static void FatalError(void* raw_args) {
+  auto called_function = reinterpret_cast<PackedArguments<void, uintptr_t>*>(raw_args)->a0;
+  fprintf(stderr, "FATAL: Called unknown CUDA function at address %p\n", reinterpret_cast<void*>(called_function));
+  __builtin_trap();
+}
+
+static void* MakeGuestCallable(const char* origin, void* func, const char* name) {
+  auto It = HostPtrInvokers.find(name);
+  if (It == HostPtrInvokers.end()) {
+    fprintf(stderr, "%s: Unknown cuda function at address %p: %s\n", origin, func, name);
+    if (stub_unknown_functions) {
+      const auto StubHostPtrInvoker = CallHostFunction<FatalError, void>;
+      LinkAddressToFunction((uintptr_t)func, reinterpret_cast<uintptr_t>(StubHostPtrInvoker));
+      return func;
+    }
+    return nullptr;
+  }
+  LinkAddressToFunction((uintptr_t)func, It->second);
+  return func;
+}
+
+struct override_entry {
+  std::string_view name;
+  void* ptr;
+};
+
+constexpr static std::array<override_entry, 2> proc_override = {
+  {{"cuGetProcAddress", (void*)cuGetProcAddress_v2}, {"cuGetProcAddress_v2", (void*)cuGetProcAddress_v2}}};
+
+CUresult cuGetProcAddress_v2(const char* symbol, void** pfn, int cudaVersion, cuuint64_t flags, CUdriverProcAddressQueryResult* symbolStatus) {
+
+  for (auto& over : proc_override) {
+    if (symbol == over.name) {
+      *pfn = over.ptr;
+      if (symbolStatus) {
+        *symbolStatus = (CUdriverProcAddressQueryResult)0;
+      }
+      return (CUresult)0; // CUDA_SUCCESS
+    }
+  }
+
+  void* ptr {};
+  auto Ret = fexfn_pack_cuGetProcAddress_v2(symbol, &ptr, cudaVersion, flags, symbolStatus);
+
+  if (!Ret) {
+    *pfn = MakeGuestCallable(__FUNCTION__, ptr, symbol);
+  }
+  return Ret;
+}
+}
+
+LOAD_LIB(libcuda)
diff --git a/ThunkLibs/libcuda/libcuda_Host.cpp b/ThunkLibs/libcuda/libcuda_Host.cpp
new file mode 100644
index 000000000..2ba10f675
--- /dev/null
+++ b/ThunkLibs/libcuda/libcuda_Host.cpp
@@ -0,0 +1,81 @@
+// SPDX-License-Identifier: MIT
+#include "common/Host.h"
+#include "cuda_defines.h"
+
+#include <stdio.h>
+#include <dlfcn.h>
+#include <string>
+#include <sys/mman.h>
+#include <unistd.h>
+#include "thunkgen_host_libcuda.inl"
+
+#define FEXFN_IMPL(fn) fexfn_impl_libcuda_##fn
+#define LDR_PTR(fn) fexldr_ptr_libcuda_##fn
+
+struct override_entry {
+  std::string_view name;
+  void* ptr;
+};
+
+constexpr static std::array<override_entry, 2> proc_override = {
+  {{"cuCtxCreate_v2", (void*)FEXFN_IMPL(cuCtxCreate_v2)}, {"cuGetExportTable", (void*)FEXFN_IMPL(cuGetExportTable)}}};
+
+static CUresult FEXFN_IMPL(cuGetProcAddress_v2)(const char* symbol, guest_layout<void**> pfn, int cudaVersion, cuuint64_t flags,
+                                                CUdriverProcAddressQueryResult* symbolStatus) {
+  host_layout<void*> host_data {};
+  void* ptr {};
+  CUresult ret {};
+  for (auto& over : proc_override) {
+    if (symbol == over.name) {
+      ptr = over.ptr;
+      if (symbolStatus) {
+        *symbolStatus = (CUdriverProcAddressQueryResult)0; // CU_GET_PROC_ADDRESS_SUCCESS
+      }
+      ret = (CUresult)0; // CUDA_SUCCESS
+      break;
+    }
+  }
+
+  if (!ptr) {
+    ret = LDR_PTR(cuGetProcAddress_v2)(symbol, &ptr, cudaVersion, flags, symbolStatus);
+  }
+
+  host_data.data = ptr;
+  *pfn.get_pointer() = to_guest(host_data);
+  return ret;
+}
+
+CUresult FEXFN_IMPL(cuCtxCreate_v2)(guest_layout<CUcontext*> pctx, unsigned int flags, CUdevice dev) {
+  host_layout<CUcontext> host_data {};
+  CUcontext ctx;
+  auto ret = LDR_PTR(cuCtxCreate_v2)(&ctx, flags, dev);
+  host_data.data = ctx;
+  *pctx.get_pointer() = to_guest(host_data);
+  return ret;
+}
+
+CUresult FEXFN_IMPL(cuGetExportTable)(guest_layout<const void**> ppExportTable, const CUuuid* pExportTableId) {
+  // This function returns a pointer to an driver internal export table that is undocumented publicly.
+  //
+  // Some documentation about the UUIDs that has been reversed.
+  // https://github.com/vosen/ZLUDA/blob/1b9ba2b2333746c5e2b05a2bf24fa6ec3828dcdf/zluda_dark_api/src/lib.rs#L197
+  //
+  // These two UUIDs are required by the static cuda runtime at startup. Might require more but it currently halts without these
+  // implemented. UUID: [0x6b, 0xd5, 0xfb, 0x6c, 0x5b, 0xf4, 0xe7, 0x4a, 0x89, 0x87, 0xd9, 0x39, 0x12, 0xfd, 0x9d, 0xf9]
+  // - cudart interface (0x68 bytes?)
+  // UUID: [0xa0, 0x94, 0x79, 0x8c, 0x2e, 0x74, 0x2e, 0x74, 0x93, 0xf2, 0x08, 0x00, 0x20, 0x0c, 0x0a, 0x66]
+  // - Tools runtime callback hooks (0x38 bytes?)
+
+  std::string uuid {};
+  for (size_t i = 0; i < sizeof(CUuuid); ++i) {
+    const bool last = (i + 1) == sizeof(CUuuid);
+    char tmp[8];
+    auto size = snprintf(tmp, 8, "0x%02x%s", pExportTableId->_0[i], last ? "" : ", ");
+    uuid += std::string_view(tmp, size);
+  }
+  fprintf(stderr, "cuGetExportTable not implemented\n");
+  fprintf(stderr, "UUID: %s\n", uuid.c_str());
+  return (CUresult)4; // CUDA_ERROR_DEINITIALIZED
+}
+
+EXPORTS(libcuda)
diff --git a/ThunkLibs/libcuda/libcuda_interface.cpp b/ThunkLibs/libcuda/libcuda_interface.cpp
new file mode 100644
index 000000000..57d4c67c4
--- /dev/null
+++ b/ThunkLibs/libcuda/libcuda_interface.cpp
@@ -0,0 +1,1571 @@
+// SPDX-License-Identifier: MIT
+// Generated file. Be careful if modifying
+#include <common/GeneratorInterface.h>
+#include "cuda_defines.h"
+template<auto>
+struct fex_gen_config {
+  unsigned version = 1;
+};
+
+template<typename>
+struct fex_gen_type {};
+
+// Type definitions
+template<>
+struct fex_gen_type<CUctx_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUmod_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUfunc_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUlib_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUkern_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUarray_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUmipmappedArray_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUtexref_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUsurfref_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUevent_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUstream_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraphicsResource_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUextMemory_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUextSemaphore_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraph_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraphNode_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraphExec_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUmemPoolHandle_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUuserObject_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraphDeviceUpdatableNode_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUasyncCallbackEntry_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUuuid_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUmemFabricHandle_st> {};
+template<>
+struct fex_gen_type<CUipcEventHandle_st> {};
+template<>
+struct fex_gen_type<CUipcMemHandle_st> {};
+template<>
+struct fex_gen_type<CUipcMem_flags_enum> {};
+template<>
+struct fex_gen_type<CUmemAttach_flags_enum> {};
+template<>
+struct fex_gen_type<CUctx_flags_enum> {};
+template<>
+struct fex_gen_type<CUevent_sched_flags_enum> {};
+template<>
+struct fex_gen_type<CUstream_flags_enum> {};
+template<>
+struct fex_gen_type<CUevent_flags_enum> {};
+template<>
+struct fex_gen_type<CUevent_record_flags_enum> {};
+template<>
+struct fex_gen_type<CUevent_wait_flags_enum> {};
+template<>
+struct fex_gen_type<CUstreamWaitValue_flags_enum> {};
+template<>
+struct fex_gen_type<CUstreamWriteValue_flags_enum> {};
+template<>
+struct fex_gen_type<CUstreamBatchMemOpType_enum> {};
+template<>
+struct fex_gen_type<CUstreamMemoryBarrier_flags_enum> {};
+template<>
+struct fex_gen_type<CUstreamBatchMemOpParams_union> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_BATCH_MEM_OP_NODE_PARAMS_v1_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_BATCH_MEM_OP_NODE_PARAMS_v2_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUoccupancy_flags_enum> {};
+template<>
+struct fex_gen_type<CUstreamUpdateCaptureDependencies_flags_enum> {};
+template<>
+struct fex_gen_type<CUasyncNotificationType_enum> {};
+template<>
+struct fex_gen_type<CUasyncNotificationInfo_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUarray_format_enum> {};
+template<>
+struct fex_gen_type<CUaddress_mode_enum> {};
+template<>
+struct fex_gen_type<CUfilter_mode_enum> {};
+template<>
+struct fex_gen_type<CUdevice_attribute_enum> {};
+template<>
+struct fex_gen_type<CUdevprop_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUpointer_attribute_enum> {};
+template<>
+struct fex_gen_type<CUfunction_attribute_enum> {};
+template<>
+struct fex_gen_type<CUfunc_cache_enum> {};
+template<>
+struct fex_gen_type<CUsharedconfig_enum> {};
+template<>
+struct fex_gen_type<CUshared_carveout_enum> {};
+template<>
+struct fex_gen_type<CUmemorytype_enum> {};
+template<>
+struct fex_gen_type<CUcomputemode_enum> {};
+template<>
+struct fex_gen_type<CUmem_advise_enum> {};
+template<>
+struct fex_gen_type<CUmem_range_attribute_enum> {};
+template<>
+struct fex_gen_type<CUjit_option_enum> {};
+template<>
+struct fex_gen_type<CUjit_target_enum> {};
+template<>
+struct fex_gen_type<CUjit_fallback_enum> {};
+template<>
+struct fex_gen_type<CUjit_cacheMode_enum> {};
+template<>
+struct fex_gen_type<CUjitInputType_enum> {};
+template<>
+struct fex_gen_type<CUlinkState_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraphicsRegisterFlags_enum> {};
+template<>
+struct fex_gen_type<CUgraphicsMapResourceFlags_enum> {};
+template<>
+struct fex_gen_type<CUarray_cubemap_face_enum> {};
+template<>
+struct fex_gen_type<CUlimit_enum> {};
+template<>
+struct fex_gen_type<CUresourcetype_enum> {};
+template<>
+struct fex_gen_type<CUaccessProperty_enum> {};
+template<>
+struct fex_gen_type<CUaccessPolicyWindow_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_KERNEL_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_KERNEL_NODE_PARAMS_v2_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_KERNEL_NODE_PARAMS_v3_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_MEMSET_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_MEMSET_NODE_PARAMS_v2_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_HOST_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_HOST_NODE_PARAMS_v2_st> {};
+template<>
+struct fex_gen_type<CUgraphConditionalNodeType_enum> {};
+template<>
+struct fex_gen_type<CUgraphNodeType_enum> {};
+template<>
+struct fex_gen_type<CUgraphDependencyType_enum> {};
+template<>
+struct fex_gen_type<CUgraphEdgeData_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraphInstantiateResult_enum> {};
+template<>
+struct fex_gen_type<CUDA_GRAPH_INSTANTIATE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUsynchronizationPolicy_enum> {};
+template<>
+struct fex_gen_type<CUclusterSchedulingPolicy_enum> {};
+template<>
+struct fex_gen_type<CUlaunchMemSyncDomain_enum> {};
+template<>
+struct fex_gen_type<CUlaunchMemSyncDomainMap_st> {};
+template<>
+struct fex_gen_type<CUlaunchAttributeID_enum> {};
+template<>
+struct fex_gen_type<CUlaunchAttributeValue_union> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUlaunchAttribute_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUlaunchConfig_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUstreamCaptureStatus_enum> {};
+template<>
+struct fex_gen_type<CUstreamCaptureMode_enum> {};
+template<>
+struct fex_gen_type<CUdriverProcAddress_flags_enum> {};
+template<>
+struct fex_gen_type<CUdriverProcAddressQueryResult_enum> {};
+template<>
+struct fex_gen_type<CUexecAffinityType_enum> {};
+template<>
+struct fex_gen_type<CUexecAffinitySmCount_st> {};
+template<>
+struct fex_gen_type<CUexecAffinityParam_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUlibraryOption_enum> {};
+template<>
+struct fex_gen_type<CUlibraryHostUniversalFunctionAndDataTable_st> {};
+template<>
+struct fex_gen_type<CUdevice_P2PAttribute_enum> {};
+template<>
+struct fex_gen_type<CUDA_MEMCPY2D_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_MEMCPY3D_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_MEMCPY3D_PEER_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_MEMCPY_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_ARRAY_DESCRIPTOR_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_ARRAY3D_DESCRIPTOR_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_ARRAY_SPARSE_PROPERTIES_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_ARRAY_MEMORY_REQUIREMENTS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_RESOURCE_DESC_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_TEXTURE_DESC_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUresourceViewFormat_enum> {};
+template<>
+struct fex_gen_type<CUDA_RESOURCE_VIEW_DESC_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUtensorMap_st> {};
+template<>
+struct fex_gen_type<CUtensorMapDataType_enum> {};
+template<>
+struct fex_gen_type<CUtensorMapInterleave_enum> {};
+template<>
+struct fex_gen_type<CUtensorMapSwizzle_enum> {};
+template<>
+struct fex_gen_type<CUtensorMapL2promotion_enum> {};
+template<>
+struct fex_gen_type<CUtensorMapFloatOOBfill_enum> {};
+template<>
+struct fex_gen_type<CUDA_POINTER_ATTRIBUTE_P2P_TOKENS_st> {};
+template<>
+struct fex_gen_type<CUDA_POINTER_ATTRIBUTE_ACCESS_FLAGS_enum> {};
+template<>
+struct fex_gen_type<CUDA_LAUNCH_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUexternalMemoryHandleType_enum> {};
+template<>
+struct fex_gen_type<CUDA_EXTERNAL_MEMORY_HANDLE_DESC_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_EXTERNAL_MEMORY_BUFFER_DESC_st> {};
+template<>
+struct fex_gen_type<CUDA_EXTERNAL_MEMORY_MIPMAPPED_ARRAY_DESC_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUexternalSemaphoreHandleType_enum> {};
+template<>
+struct fex_gen_type<CUDA_EXTERNAL_SEMAPHORE_HANDLE_DESC_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_EXTERNAL_SEMAPHORE_SIGNAL_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_EXTERNAL_SEMAPHORE_WAIT_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_EXT_SEM_SIGNAL_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_EXT_SEM_SIGNAL_NODE_PARAMS_v2_st> {};
+template<>
+struct fex_gen_type<CUDA_EXT_SEM_WAIT_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_EXT_SEM_WAIT_NODE_PARAMS_v2_st> {};
+template<>
+struct fex_gen_type<CUmemAllocationHandleType_enum> {};
+template<>
+struct fex_gen_type<CUmemAccess_flags_enum> {};
+template<>
+struct fex_gen_type<CUmemLocationType_enum> {};
+template<>
+struct fex_gen_type<CUmemAllocationType_enum> {};
+template<>
+struct fex_gen_type<CUmemAllocationGranularity_flags_enum> {};
+template<>
+struct fex_gen_type<CUmemRangeHandleType_enum> {};
+template<>
+struct fex_gen_type<CUarraySparseSubresourceType_enum> {};
+template<>
+struct fex_gen_type<CUmemOperationType_enum> {};
+template<>
+struct fex_gen_type<CUmemHandleType_enum> {};
+template<>
+struct fex_gen_type<CUarrayMapInfo_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUmemLocation_st> {};
+template<>
+struct fex_gen_type<CUmemAllocationCompType_enum> {};
+template<>
+struct fex_gen_type<CUmemAllocationProp_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUmulticastGranularity_flags_enum> {};
+template<>
+struct fex_gen_type<CUmulticastObjectProp_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUmemAccessDesc_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraphExecUpdateResult_enum> {};
+template<>
+struct fex_gen_type<CUgraphExecUpdateResultInfo_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUmemPool_attribute_enum> {};
+template<>
+struct fex_gen_type<CUmemPoolProps_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUmemPoolPtrExportData_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_MEM_ALLOC_NODE_PARAMS_v1_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_MEM_ALLOC_NODE_PARAMS_v2_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_MEM_FREE_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraphMem_attribute_enum> {};
+template<>
+struct fex_gen_type<CUDA_CHILD_GRAPH_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_EVENT_RECORD_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUDA_EVENT_WAIT_NODE_PARAMS_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUgraphNodeParams_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUflushGPUDirectRDMAWritesOptions_enum> {};
+template<>
+struct fex_gen_type<CUGPUDirectRDMAWritesOrdering_enum> {};
+template<>
+struct fex_gen_type<CUflushGPUDirectRDMAWritesScope_enum> {};
+template<>
+struct fex_gen_type<CUflushGPUDirectRDMAWritesTarget_enum> {};
+template<>
+struct fex_gen_type<CUgraphDebugDot_flags_enum> {};
+template<>
+struct fex_gen_type<CUuserObject_flags_enum> {};
+template<>
+struct fex_gen_type<CUuserObjectRetain_flags_enum> {};
+template<>
+struct fex_gen_type<CUgraphInstantiate_flags_enum> {};
+template<>
+struct fex_gen_type<CUdeviceNumaConfig_enum> {};
+template<>
+struct fex_gen_type<CUmoduleLoadingMode_enum> {};
+template<>
+struct fex_gen_type<CUfunctionLoadingState_enum> {};
+template<>
+struct fex_gen_type<CUcoredumpSettings_enum> {};
+template<>
+struct fex_gen_type<CUgreenCtx_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUdevResourceDesc_st> : fexgen::opaque_type {};
+template<>
+struct fex_gen_type<CUdevSmResource_st> {};
+template<>
+struct fex_gen_type<CUdevResource_st> : fexgen::opaque_type {};
+
+// Function definitions
+namespace internal {
+template<auto>
+struct fex_gen_config : fexgen::generate_guest_symtable, fexgen::indirect_guest_calls {};
+
+// Function, parameter index, parameter type [optional]
+template<auto, int, typename = void>
+struct fex_gen_param {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGetErrorString> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGetErrorName> {};
+#endif
+template<>
+struct fex_gen_config<cuInit> {};
+template<>
+struct fex_gen_config<cuDriverGetVersion> {};
+template<>
+struct fex_gen_config<cuDeviceGet> {};
+template<>
+struct fex_gen_config<cuDeviceGetCount> {};
+template<>
+struct fex_gen_config<cuDeviceGetName> {};
+template<>
+struct fex_gen_config<cuDeviceGetUuid> {};
+template<>
+struct fex_gen_config<cuDeviceGetUuid_v2> {};
+template<>
+struct fex_gen_config<cuDeviceGetLuid> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuDeviceTotalMem_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuDeviceGetTexture1DLinearMaxWidth> {};
+#endif
+template<>
+struct fex_gen_config<cuDeviceGetAttribute> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuDeviceGetNvSciSyncAttributes> {};
+#endif
+template<>
+struct fex_gen_config<cuDeviceSetMemPool> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuDeviceGetMemPool> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuDeviceGetDefaultMemPool> {};
+#endif
+template<>
+struct fex_gen_config<cuDeviceGetExecAffinitySupport> {};
+template<>
+struct fex_gen_config<cuFlushGPUDirectRDMAWrites> {};
+template<>
+struct fex_gen_config<cuDeviceGetProperties> {};
+template<>
+struct fex_gen_config<cuDeviceComputeCapability> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuDevicePrimaryCtxRetain> {};
+#endif
+template<>
+struct fex_gen_config<cuDevicePrimaryCtxRelease_v2> {};
+template<>
+struct fex_gen_config<cuDevicePrimaryCtxSetFlags_v2> {};
+template<>
+struct fex_gen_config<cuDevicePrimaryCtxGetState> {};
+template<>
+struct fex_gen_config<cuDevicePrimaryCtxReset_v2> {};
+template<>
+struct fex_gen_config<cuCtxCreate_v2> : fexgen::custom_host_impl {};
+template<>
+struct fex_gen_param<cuCtxCreate_v2, 0, CUcontext*> : fexgen::ptr_passthrough {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCtxCreate_v3> {};
+#endif
+template<>
+struct fex_gen_config<cuCtxDestroy_v2> {};
+template<>
+struct fex_gen_config<cuCtxPushCurrent_v2> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCtxPopCurrent_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuCtxSetCurrent> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCtxGetCurrent> {};
+#endif
+template<>
+struct fex_gen_config<cuCtxGetDevice> {};
+template<>
+struct fex_gen_config<cuCtxGetFlags> {};
+template<>
+struct fex_gen_config<cuCtxSetFlags> {};
+template<>
+struct fex_gen_config<cuCtxGetId> {};
+template<>
+struct fex_gen_config<cuCtxSynchronize> {};
+template<>
+struct fex_gen_config<cuCtxSetLimit> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCtxGetLimit> {};
+#endif
+template<>
+struct fex_gen_config<cuCtxGetCacheConfig> {};
+template<>
+struct fex_gen_config<cuCtxSetCacheConfig> {};
+template<>
+struct fex_gen_config<cuCtxGetApiVersion> {};
+template<>
+struct fex_gen_config<cuCtxGetStreamPriorityRange> {};
+template<>
+struct fex_gen_config<cuCtxResetPersistingL2Cache> {};
+template<>
+struct fex_gen_config<cuCtxGetExecAffinity> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCtxAttach> {};
+#endif
+template<>
+struct fex_gen_config<cuCtxDetach> {};
+template<>
+struct fex_gen_config<cuCtxGetSharedMemConfig> {};
+template<>
+struct fex_gen_config<cuCtxSetSharedMemConfig> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuModuleLoad> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuModuleLoadData> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuModuleLoadDataEx> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuModuleLoadFatBinary> {};
+#endif
+template<>
+struct fex_gen_config<cuModuleUnload> {};
+template<>
+struct fex_gen_config<cuModuleGetLoadingMode> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuModuleGetFunction> {};
+#endif
+template<>
+struct fex_gen_config<cuModuleGetFunctionCount> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuModuleEnumerateFunctions> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuModuleGetGlobal_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLinkCreate_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLinkAddData_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLinkAddFile_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLinkComplete> {};
+#endif
+template<>
+struct fex_gen_config<cuLinkDestroy> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuModuleGetTexRef> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuModuleGetSurfRef> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLibraryLoadData> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLibraryLoadFromFile> {};
+#endif
+template<>
+struct fex_gen_config<cuLibraryUnload> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLibraryGetKernel> {};
+#endif
+template<>
+struct fex_gen_config<cuLibraryGetKernelCount> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLibraryEnumerateKernels> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLibraryGetModule> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuKernelGetFunction> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLibraryGetGlobal> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLibraryGetManaged> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLibraryGetUnifiedFunction> {};
+#endif
+template<>
+struct fex_gen_config<cuKernelGetAttribute> {};
+template<>
+struct fex_gen_config<cuKernelSetAttribute> {};
+template<>
+struct fex_gen_config<cuKernelSetCacheConfig> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuKernelGetName> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuKernelGetParamInfo> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemGetInfo_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuMemAlloc_v2> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemAllocPitch_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuMemFree_v2> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemGetAddressRange_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemAllocHost_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemFreeHost> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemHostAlloc> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemHostGetDevicePointer_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemHostGetFlags> {};
+#endif
+template<>
+struct fex_gen_config<cuMemAllocManaged> {};
+#ifndef IS_32BIT_THUNK
+// Disabled due to callbacks
+// template<>
+// struct fex_gen_config<cuDeviceRegisterAsyncNotification>  {};
+#endif
+template<>
+struct fex_gen_config<cuDeviceUnregisterAsyncNotification> {};
+template<>
+struct fex_gen_config<cuDeviceGetByPCIBusId> {};
+template<>
+struct fex_gen_config<cuDeviceGetPCIBusId> {};
+template<>
+struct fex_gen_config<cuIpcGetEventHandle> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuIpcOpenEventHandle> {};
+#endif
+template<>
+struct fex_gen_config<cuIpcGetMemHandle> {};
+template<>
+struct fex_gen_config<cuIpcOpenMemHandle_v2> {};
+template<>
+struct fex_gen_config<cuIpcCloseMemHandle> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemHostRegister_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemHostUnregister> {};
+#endif
+template<>
+struct fex_gen_config<cuMemcpy> {};
+template<>
+struct fex_gen_config<cuMemcpyPeer> {};
+template<>
+struct fex_gen_config<cuMemcpyHtoD_v2> {};
+template<>
+struct fex_gen_param<cuMemcpyHtoD_v2, 1, const void*> : fexgen::assume_compatible_data_layout {};
+template<>
+struct fex_gen_config<cuMemcpyDtoH_v2> {};
+template<>
+struct fex_gen_param<cuMemcpyDtoH_v2, 0, void*> : fexgen::assume_compatible_data_layout {};
+template<>
+struct fex_gen_config<cuMemcpyDtoD_v2> {};
+template<>
+struct fex_gen_config<cuMemcpyDtoA_v2> {};
+template<>
+struct fex_gen_config<cuMemcpyAtoD_v2> {};
+template<>
+struct fex_gen_config<cuMemcpyHtoA_v2> {};
+template<>
+struct fex_gen_param<cuMemcpyHtoA_v2, 2, const void*> : fexgen::assume_compatible_data_layout {};
+template<>
+struct fex_gen_config<cuMemcpyAtoH_v2> {};
+template<>
+struct fex_gen_param<cuMemcpyAtoH_v2, 0, void*> : fexgen::assume_compatible_data_layout {};
+template<>
+struct fex_gen_config<cuMemcpyAtoA_v2> {};
+template<>
+struct fex_gen_config<cuMemcpy2D_v2> {};
+template<>
+struct fex_gen_config<cuMemcpy2DUnaligned_v2> {};
+template<>
+struct fex_gen_config<cuMemcpy3D_v2> {};
+template<>
+struct fex_gen_config<cuMemcpy3DPeer> {};
+template<>
+struct fex_gen_config<cuMemcpyAsync> {};
+template<>
+struct fex_gen_config<cuMemcpyPeerAsync> {};
+template<>
+struct fex_gen_config<cuMemcpyHtoDAsync_v2> {};
+template<>
+struct fex_gen_param<cuMemcpyHtoDAsync_v2, 1, const void*> : fexgen::assume_compatible_data_layout {};
+template<>
+struct fex_gen_config<cuMemcpyDtoHAsync_v2> {};
+template<>
+struct fex_gen_param<cuMemcpyDtoHAsync_v2, 0, void*> : fexgen::assume_compatible_data_layout {};
+template<>
+struct fex_gen_config<cuMemcpyDtoDAsync_v2> {};
+template<>
+struct fex_gen_config<cuMemcpyHtoAAsync_v2> {};
+template<>
+struct fex_gen_param<cuMemcpyHtoAAsync_v2, 2, const void*> : fexgen::assume_compatible_data_layout {};
+template<>
+struct fex_gen_config<cuMemcpyAtoHAsync_v2> {};
+template<>
+struct fex_gen_param<cuMemcpyAtoHAsync_v2, 0, void*> : fexgen::assume_compatible_data_layout {};
+template<>
+struct fex_gen_config<cuMemcpy2DAsync_v2> {};
+template<>
+struct fex_gen_config<cuMemcpy3DAsync_v2> {};
+template<>
+struct fex_gen_config<cuMemcpy3DPeerAsync> {};
+template<>
+struct fex_gen_config<cuMemsetD8_v2> {};
+template<>
+struct fex_gen_config<cuMemsetD16_v2> {};
+template<>
+struct fex_gen_config<cuMemsetD32_v2> {};
+template<>
+struct fex_gen_config<cuMemsetD2D8_v2> {};
+template<>
+struct fex_gen_config<cuMemsetD2D16_v2> {};
+template<>
+struct fex_gen_config<cuMemsetD2D32_v2> {};
+template<>
+struct fex_gen_config<cuMemsetD8Async> {};
+template<>
+struct fex_gen_config<cuMemsetD16Async> {};
+template<>
+struct fex_gen_config<cuMemsetD32Async> {};
+template<>
+struct fex_gen_config<cuMemsetD2D8Async> {};
+template<>
+struct fex_gen_config<cuMemsetD2D16Async> {};
+template<>
+struct fex_gen_config<cuMemsetD2D32Async> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuArrayCreate_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuArrayGetDescriptor_v2> {};
+template<>
+struct fex_gen_config<cuArrayGetSparseProperties> {};
+template<>
+struct fex_gen_config<cuMipmappedArrayGetSparseProperties> {};
+template<>
+struct fex_gen_config<cuArrayGetMemoryRequirements> {};
+template<>
+struct fex_gen_config<cuMipmappedArrayGetMemoryRequirements> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuArrayGetPlane> {};
+#endif
+template<>
+struct fex_gen_config<cuArrayDestroy> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuArray3DCreate_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuArray3DGetDescriptor_v2> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMipmappedArrayCreate> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMipmappedArrayGetLevel> {};
+#endif
+template<>
+struct fex_gen_config<cuMipmappedArrayDestroy> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemGetHandleForAddressRange> {};
+#endif
+template<>
+struct fex_gen_config<cuMemAddressReserve> {};
+template<>
+struct fex_gen_config<cuMemAddressFree> {};
+template<>
+struct fex_gen_config<cuMemCreate> {};
+template<>
+struct fex_gen_config<cuMemRelease> {};
+template<>
+struct fex_gen_config<cuMemMap> {};
+template<>
+struct fex_gen_config<cuMemMapArrayAsync> {};
+template<>
+struct fex_gen_config<cuMemUnmap> {};
+template<>
+struct fex_gen_config<cuMemSetAccess> {};
+template<>
+struct fex_gen_config<cuMemGetAccess> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemExportToShareableHandle> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemImportFromShareableHandle> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemGetAllocationGranularity> {};
+#endif
+template<>
+struct fex_gen_config<cuMemGetAllocationPropertiesFromHandle> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemRetainAllocationHandle> {};
+#endif
+template<>
+struct fex_gen_config<cuMemFreeAsync> {};
+template<>
+struct fex_gen_config<cuMemAllocAsync> {};
+template<>
+struct fex_gen_config<cuMemPoolTrimTo> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemPoolSetAttribute> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemPoolGetAttribute> {};
+#endif
+template<>
+struct fex_gen_config<cuMemPoolSetAccess> {};
+template<>
+struct fex_gen_config<cuMemPoolGetAccess> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemPoolCreate> {};
+#endif
+template<>
+struct fex_gen_config<cuMemPoolDestroy> {};
+template<>
+struct fex_gen_config<cuMemAllocFromPoolAsync> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemPoolExportToShareableHandle> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemPoolImportFromShareableHandle> {};
+#endif
+template<>
+struct fex_gen_config<cuMemPoolExportPointer> {};
+template<>
+struct fex_gen_config<cuMemPoolImportPointer> {};
+template<>
+struct fex_gen_config<cuMulticastCreate> {};
+template<>
+struct fex_gen_config<cuMulticastAddDevice> {};
+template<>
+struct fex_gen_config<cuMulticastBindMem> {};
+template<>
+struct fex_gen_config<cuMulticastBindAddr> {};
+template<>
+struct fex_gen_config<cuMulticastUnbind> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMulticastGetGranularity> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuPointerGetAttribute> {};
+#endif
+template<>
+struct fex_gen_config<cuMemPrefetchAsync> {};
+template<>
+struct fex_gen_config<cuMemPrefetchAsync_v2> {};
+template<>
+struct fex_gen_config<cuMemAdvise> {};
+template<>
+struct fex_gen_config<cuMemAdvise_v2> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemRangeGetAttribute> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuMemRangeGetAttributes> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuPointerSetAttribute> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuPointerGetAttributes> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuStreamCreate> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuStreamCreateWithPriority> {};
+#endif
+template<>
+struct fex_gen_config<cuStreamGetPriority> {};
+template<>
+struct fex_gen_config<cuStreamGetFlags> {};
+template<>
+struct fex_gen_config<cuStreamGetId> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuStreamGetCtx> {};
+#endif
+template<>
+struct fex_gen_config<cuStreamWaitEvent> {};
+// Disabled due to callbacks
+// template<>
+// struct fex_gen_config<cuStreamAddCallback>  {};
+template<>
+struct fex_gen_config<cuStreamBeginCapture_v2> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuStreamBeginCaptureToGraph> {};
+#endif
+template<>
+struct fex_gen_config<cuThreadExchangeStreamCaptureMode> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuStreamEndCapture> {};
+#endif
+template<>
+struct fex_gen_config<cuStreamIsCapturing> {};
+#ifndef IS_32BIT_THUNK
+// template<>
+// struct fex_gen_config<cuStreamGetCaptureInfo_v2>  {};
+#endif
+#ifndef IS_32BIT_THUNK
+// template<>
+// struct fex_gen_config<cuStreamGetCaptureInfo_v3>  {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuStreamUpdateCaptureDependencies> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuStreamUpdateCaptureDependencies_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuStreamAttachMemAsync> {};
+template<>
+struct fex_gen_config<cuStreamQuery> {};
+template<>
+struct fex_gen_config<cuStreamSynchronize> {};
+template<>
+struct fex_gen_config<cuStreamDestroy_v2> {};
+template<>
+struct fex_gen_config<cuStreamCopyAttributes> {};
+template<>
+struct fex_gen_config<cuStreamGetAttribute> {};
+template<>
+struct fex_gen_config<cuStreamSetAttribute> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuEventCreate> {};
+#endif
+template<>
+struct fex_gen_config<cuEventRecord> {};
+template<>
+struct fex_gen_config<cuEventRecordWithFlags> {};
+template<>
+struct fex_gen_config<cuEventQuery> {};
+template<>
+struct fex_gen_config<cuEventSynchronize> {};
+template<>
+struct fex_gen_config<cuEventDestroy_v2> {};
+template<>
+struct fex_gen_config<cuEventElapsedTime> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuImportExternalMemory> {};
+#endif
+template<>
+struct fex_gen_config<cuExternalMemoryGetMappedBuffer> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuExternalMemoryGetMappedMipmappedArray> {};
+#endif
+template<>
+struct fex_gen_config<cuDestroyExternalMemory> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuImportExternalSemaphore> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuSignalExternalSemaphoresAsync> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuWaitExternalSemaphoresAsync> {};
+#endif
+template<>
+struct fex_gen_config<cuDestroyExternalSemaphore> {};
+template<>
+struct fex_gen_config<cuStreamWaitValue32_v2> {};
+template<>
+struct fex_gen_config<cuStreamWaitValue64_v2> {};
+template<>
+struct fex_gen_config<cuStreamWriteValue32_v2> {};
+template<>
+struct fex_gen_config<cuStreamWriteValue64_v2> {};
+template<>
+struct fex_gen_config<cuStreamBatchMemOp_v2> {};
+template<>
+struct fex_gen_config<cuFuncGetAttribute> {};
+template<>
+struct fex_gen_config<cuFuncSetAttribute> {};
+template<>
+struct fex_gen_config<cuFuncSetCacheConfig> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuFuncGetModule> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuFuncGetName> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuFuncGetParamInfo> {};
+#endif
+template<>
+struct fex_gen_config<cuFuncIsLoaded> {};
+template<>
+struct fex_gen_config<cuFuncLoad> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLaunchKernel> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLaunchKernelEx> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLaunchCooperativeKernel> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLaunchCooperativeKernelMultiDevice> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuLaunchHostFunc> {};
+#endif
+template<>
+struct fex_gen_config<cuFuncSetBlockShape> {};
+template<>
+struct fex_gen_config<cuFuncSetSharedSize> {};
+template<>
+struct fex_gen_config<cuParamSetSize> {};
+template<>
+struct fex_gen_config<cuParamSeti> {};
+template<>
+struct fex_gen_config<cuParamSetf> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuParamSetv> {};
+#endif
+template<>
+struct fex_gen_config<cuLaunch> {};
+template<>
+struct fex_gen_config<cuLaunchGrid> {};
+template<>
+struct fex_gen_config<cuLaunchGridAsync> {};
+template<>
+struct fex_gen_config<cuParamSetTexRef> {};
+template<>
+struct fex_gen_config<cuFuncSetSharedMemConfig> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphCreate> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddKernelNode_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphKernelNodeGetParams_v2> {};
+template<>
+struct fex_gen_config<cuGraphKernelNodeSetParams_v2> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddMemcpyNode> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphMemcpyNodeGetParams> {};
+template<>
+struct fex_gen_config<cuGraphMemcpyNodeSetParams> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddMemsetNode> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphMemsetNodeGetParams> {};
+template<>
+struct fex_gen_config<cuGraphMemsetNodeSetParams> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddHostNode> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphHostNodeGetParams> {};
+template<>
+struct fex_gen_config<cuGraphHostNodeSetParams> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddChildGraphNode> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphChildGraphNodeGetGraph> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddEmptyNode> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddEventRecordNode> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphEventRecordNodeGetEvent> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphEventRecordNodeSetEvent> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddEventWaitNode> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphEventWaitNodeGetEvent> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphEventWaitNodeSetEvent> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddExternalSemaphoresSignalNode> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphExternalSemaphoresSignalNodeGetParams> {};
+template<>
+struct fex_gen_config<cuGraphExternalSemaphoresSignalNodeSetParams> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddExternalSemaphoresWaitNode> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphExternalSemaphoresWaitNodeGetParams> {};
+template<>
+struct fex_gen_config<cuGraphExternalSemaphoresWaitNodeSetParams> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddBatchMemOpNode> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphBatchMemOpNodeGetParams> {};
+template<>
+struct fex_gen_config<cuGraphBatchMemOpNodeSetParams> {};
+template<>
+struct fex_gen_config<cuGraphExecBatchMemOpNodeSetParams> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddMemAllocNode> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphMemAllocNodeGetParams> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddMemFreeNode> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphMemFreeNodeGetParams> {};
+template<>
+struct fex_gen_config<cuDeviceGraphMemTrim> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuDeviceGetGraphMemAttribute> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuDeviceSetGraphMemAttribute> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphClone> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphNodeFindInClone> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphNodeGetType> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphGetNodes> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphGetRootNodes> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphGetEdges> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphGetEdges_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphNodeGetDependencies> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphNodeGetDependencies_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphNodeGetDependentNodes> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphNodeGetDependentNodes_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddDependencies> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddDependencies_v2> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphRemoveDependencies> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphRemoveDependencies_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphDestroyNode> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphInstantiateWithFlags> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphInstantiateWithParams> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphExecGetFlags> {};
+template<>
+struct fex_gen_config<cuGraphExecKernelNodeSetParams_v2> {};
+template<>
+struct fex_gen_config<cuGraphExecMemcpyNodeSetParams> {};
+template<>
+struct fex_gen_config<cuGraphExecMemsetNodeSetParams> {};
+template<>
+struct fex_gen_config<cuGraphExecHostNodeSetParams> {};
+template<>
+struct fex_gen_config<cuGraphExecChildGraphNodeSetParams> {};
+template<>
+struct fex_gen_config<cuGraphExecEventRecordNodeSetEvent> {};
+template<>
+struct fex_gen_config<cuGraphExecEventWaitNodeSetEvent> {};
+template<>
+struct fex_gen_config<cuGraphExecExternalSemaphoresSignalNodeSetParams> {};
+template<>
+struct fex_gen_config<cuGraphExecExternalSemaphoresWaitNodeSetParams> {};
+template<>
+struct fex_gen_config<cuGraphNodeSetEnabled> {};
+template<>
+struct fex_gen_config<cuGraphNodeGetEnabled> {};
+template<>
+struct fex_gen_config<cuGraphUpload> {};
+template<>
+struct fex_gen_config<cuGraphLaunch> {};
+template<>
+struct fex_gen_config<cuGraphExecDestroy> {};
+template<>
+struct fex_gen_config<cuGraphDestroy> {};
+template<>
+struct fex_gen_config<cuGraphExecUpdate_v2> {};
+template<>
+struct fex_gen_config<cuGraphKernelNodeCopyAttributes> {};
+template<>
+struct fex_gen_config<cuGraphKernelNodeGetAttribute> {};
+template<>
+struct fex_gen_config<cuGraphKernelNodeSetAttribute> {};
+template<>
+struct fex_gen_config<cuGraphDebugDotPrint> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuUserObjectCreate> {};
+#endif
+template<>
+struct fex_gen_config<cuUserObjectRetain> {};
+template<>
+struct fex_gen_config<cuUserObjectRelease> {};
+template<>
+struct fex_gen_config<cuGraphRetainUserObject> {};
+template<>
+struct fex_gen_config<cuGraphReleaseUserObject> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddNode> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphAddNode_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphNodeSetParams> {};
+template<>
+struct fex_gen_config<cuGraphExecNodeSetParams> {};
+template<>
+struct fex_gen_config<cuGraphConditionalHandleCreate> {};
+template<>
+struct fex_gen_config<cuOccupancyMaxActiveBlocksPerMultiprocessor> {};
+template<>
+struct fex_gen_config<cuOccupancyMaxActiveBlocksPerMultiprocessorWithFlags> {};
+// template<>
+// struct fex_gen_config<cuOccupancyMaxPotentialBlockSize>  {};
+// template<>
+// struct fex_gen_config<cuOccupancyMaxPotentialBlockSizeWithFlags>  {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuOccupancyAvailableDynamicSMemPerBlock> {};
+#endif
+template<>
+struct fex_gen_config<cuOccupancyMaxPotentialClusterSize> {};
+template<>
+struct fex_gen_config<cuOccupancyMaxActiveClusters> {};
+template<>
+struct fex_gen_config<cuTexRefSetArray> {};
+template<>
+struct fex_gen_config<cuTexRefSetMipmappedArray> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuTexRefSetAddress_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuTexRefSetAddress2D_v3> {};
+template<>
+struct fex_gen_config<cuTexRefSetFormat> {};
+template<>
+struct fex_gen_config<cuTexRefSetAddressMode> {};
+template<>
+struct fex_gen_config<cuTexRefSetFilterMode> {};
+template<>
+struct fex_gen_config<cuTexRefSetMipmapFilterMode> {};
+template<>
+struct fex_gen_config<cuTexRefSetMipmapLevelBias> {};
+template<>
+struct fex_gen_config<cuTexRefSetMipmapLevelClamp> {};
+template<>
+struct fex_gen_config<cuTexRefSetMaxAnisotropy> {};
+template<>
+struct fex_gen_config<cuTexRefSetBorderColor> {};
+template<>
+struct fex_gen_config<cuTexRefSetFlags> {};
+template<>
+struct fex_gen_config<cuTexRefGetAddress_v2> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuTexRefGetArray> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuTexRefGetMipmappedArray> {};
+#endif
+template<>
+struct fex_gen_config<cuTexRefGetAddressMode> {};
+template<>
+struct fex_gen_config<cuTexRefGetFilterMode> {};
+template<>
+struct fex_gen_config<cuTexRefGetFormat> {};
+template<>
+struct fex_gen_config<cuTexRefGetMipmapFilterMode> {};
+template<>
+struct fex_gen_config<cuTexRefGetMipmapLevelBias> {};
+template<>
+struct fex_gen_config<cuTexRefGetMipmapLevelClamp> {};
+template<>
+struct fex_gen_config<cuTexRefGetMaxAnisotropy> {};
+template<>
+struct fex_gen_config<cuTexRefGetBorderColor> {};
+template<>
+struct fex_gen_config<cuTexRefGetFlags> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuTexRefCreate> {};
+#endif
+template<>
+struct fex_gen_config<cuTexRefDestroy> {};
+template<>
+struct fex_gen_config<cuSurfRefSetArray> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuSurfRefGetArray> {};
+#endif
+template<>
+struct fex_gen_config<cuTexObjectCreate> {};
+template<>
+struct fex_gen_config<cuTexObjectDestroy> {};
+template<>
+struct fex_gen_config<cuTexObjectGetResourceDesc> {};
+template<>
+struct fex_gen_config<cuTexObjectGetTextureDesc> {};
+template<>
+struct fex_gen_config<cuTexObjectGetResourceViewDesc> {};
+template<>
+struct fex_gen_config<cuSurfObjectCreate> {};
+template<>
+struct fex_gen_config<cuSurfObjectDestroy> {};
+template<>
+struct fex_gen_config<cuSurfObjectGetResourceDesc> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuTensorMapEncodeTiled> {};
+#endif
+// template<>
+// struct fex_gen_config<cuTensorMapEncodeIm2col>  {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuTensorMapReplaceAddress> {};
+#endif
+template<>
+struct fex_gen_config<cuDeviceCanAccessPeer> {};
+template<>
+struct fex_gen_config<cuCtxEnablePeerAccess> {};
+template<>
+struct fex_gen_config<cuCtxDisablePeerAccess> {};
+template<>
+struct fex_gen_config<cuDeviceGetP2PAttribute> {};
+template<>
+struct fex_gen_config<cuGraphicsUnregisterResource> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphicsSubResourceGetMappedArray> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphicsResourceGetMappedMipmappedArray> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphicsResourceGetMappedPointer_v2> {};
+#endif
+template<>
+struct fex_gen_config<cuGraphicsResourceSetMapFlags_v2> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphicsMapResources> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGraphicsUnmapResources> {};
+#endif
+template<>
+struct fex_gen_config<cuGetProcAddress_v2> : fexgen::custom_host_impl, fexgen::custom_guest_entrypoint, fexgen::returns_guest_pointer {};
+template<>
+struct fex_gen_param<cuGetProcAddress_v2, 1, void**> : fexgen::ptr_passthrough {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCoredumpGetAttribute> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCoredumpGetAttributeGlobal> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCoredumpSetAttribute> {};
+#endif
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCoredumpSetAttributeGlobal> {};
+#endif
+template<>
+struct fex_gen_config<cuGetExportTable> : fexgen::custom_host_impl {};
+template<>
+struct fex_gen_param<cuGetExportTable, 0, const void**> : fexgen::ptr_passthrough {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuGreenCtxCreate> {};
+#endif
+template<>
+struct fex_gen_config<cuGreenCtxDestroy> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuCtxFromGreenCtx> {};
+#endif
+template<>
+struct fex_gen_config<cuDeviceGetDevResource> {};
+template<>
+struct fex_gen_config<cuCtxGetDevResource> {};
+template<>
+struct fex_gen_config<cuGreenCtxGetDevResource> {};
+template<>
+struct fex_gen_config<cuDevSmResourceSplitByCount> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuDevResourceGenerateDesc> {};
+#endif
+template<>
+struct fex_gen_config<cuGreenCtxRecordEvent> {};
+template<>
+struct fex_gen_config<cuGreenCtxWaitEvent> {};
+#ifndef IS_32BIT_THUNK
+template<>
+struct fex_gen_config<cuStreamGetGreenCtx> {};
+#endif
+} // namespace internal