From 23dc0ba4777faeb635fbe3ae2f77dd64574d703a Mon Sep 17 00:00:00 2001
From: =?UTF-8?q?Michael=20M=C3=BCller?= <michael@fds-team.de>
Date: Sat, 3 Jan 2015 03:39:11 +0100
Subject: [PATCH] nvcuda: First implementation. (rev 2)

Changes by Sebastian Lackner <sebastian@fds-team.de>:
* Add static to a couple of functions
* TLS callbacks should be declared as CDECL
* Do not unlock the critical section while processing TLS callbacks
---
 configure.ac                  |    1 +
 dlls/nvcuda/Makefile.in       |    1 +
 dlls/nvcuda/internal.c        |  515 ++++++++++
 dlls/nvcuda/nvcuda.c          | 2097 ++++++++++++++++++++++++++++++++++++++++-
 dlls/nvcuda/nvcuda.h          |   29 +
 dlls/nvcuda/nvcuda.spec       |  492 +++++-----
 dlls/nvcuda/tests/Makefile.in |    4 +
 dlls/nvcuda/tests/nvcuda.c    |  168 ++++
 8 files changed, 3056 insertions(+), 251 deletions(-)
 create mode 100644 dlls/nvcuda/internal.c
 create mode 100644 dlls/nvcuda/nvcuda.h
 create mode 100644 dlls/nvcuda/tests/Makefile.in
 create mode 100644 dlls/nvcuda/tests/nvcuda.c

diff --git a/configure.ac b/configure.ac
index 7b6f885..55d8931 100644
--- a/configure.ac
+++ b/configure.ac
@@ -3555,6 +3555,7 @@ WINE_CONFIG_MAKEFILE(dlls/ntoskrnl.exe)
 WINE_CONFIG_MAKEFILE(dlls/ntprint)
 WINE_CONFIG_MAKEFILE(dlls/ntprint/tests)
 WINE_CONFIG_MAKEFILE(dlls/nvcuda)
+WINE_CONFIG_MAKEFILE(dlls/nvcuda/tests)
 WINE_CONFIG_MAKEFILE(dlls/objsel)
 WINE_CONFIG_MAKEFILE(dlls/odbc32)
 WINE_CONFIG_MAKEFILE(dlls/odbccp32)
diff --git a/dlls/nvcuda/Makefile.in b/dlls/nvcuda/Makefile.in
index 4b33278..0921ce6 100644
--- a/dlls/nvcuda/Makefile.in
+++ b/dlls/nvcuda/Makefile.in
@@ -1,6 +1,7 @@
 MODULE    = nvcuda.dll
 
 C_SRCS = \
+	internal.c \
 	nvcuda.c
 
 RC_SRCS = nvcuda.rc
diff --git a/dlls/nvcuda/internal.c b/dlls/nvcuda/internal.c
new file mode 100644
index 0000000..611e310
--- /dev/null
+++ b/dlls/nvcuda/internal.c
@@ -0,0 +1,515 @@
+/*
+ * Copyright (C) 2014-2015 Michael Müller
+ * Copyright (C) 2014-2015 Sebastian Lackner
+ *
+ * This library is free software; you can redistribute it and/or
+ * modify it under the terms of the GNU Lesser General Public
+ * License as published by the Free Software Foundation; either
+ * version 2.1 of the License, or (at your option) any later version.
+ *
+ * This library is distributed in the hope that it will be useful,
+ * but WITHOUT ANY WARRANTY; without even the implied warranty of
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the GNU
+ * Lesser General Public License for more details.
+ *
+ * You should have received a copy of the GNU Lesser General Public
+ * License along with this library; if not, write to the Free Software
+ * Foundation, Inc., 51 Franklin St, Fifth Floor, Boston, MA 02110-1301, USA
+ */
+
+#include "config.h"
+#include "wine/port.h"
+
+#include <stdarg.h>
+#include <stdio.h>
+
+#include "windef.h"
+#include "winbase.h"
+#include "wine/library.h"
+#include "wine/debug.h"
+#include "wine/list.h"
+#include "cuda.h"
+#include "nvcuda.h"
+
+WINE_DEFAULT_DEBUG_CHANNEL(nvcuda);
+
+struct tls_callback_entry
+{
+    struct list entry;
+    void (CDECL *callback)(DWORD, void *);
+    void *userdata;
+    ULONG count;
+};
+
+static struct list tls_callbacks = LIST_INIT( tls_callbacks );
+
+static RTL_CRITICAL_SECTION tls_callback_section;
+static RTL_CRITICAL_SECTION_DEBUG critsect_debug =
+{
+    0, 0, &tls_callback_section,
+    { &critsect_debug.ProcessLocksList, &critsect_debug.ProcessLocksList },
+      0, 0, { (DWORD_PTR)(__FILE__ ": tls_callback_section") }
+};
+static RTL_CRITICAL_SECTION tls_callback_section = { &critsect_debug, -1, 0, 0, 0, 0 };
+
+void cuda_process_tls_callbacks(DWORD reason)
+{
+    struct list *ptr;
+
+    TRACE("(%d)\n", reason);
+
+    if (reason != DLL_THREAD_DETACH)
+        return;
+
+    EnterCriticalSection( &tls_callback_section );
+    ptr = list_head( &tls_callbacks );
+    while (ptr)
+    {
+        struct tls_callback_entry *callback = LIST_ENTRY( ptr, struct tls_callback_entry, entry );
+        callback->count++;
+
+        TRACE("calling handler %p(0, %p)\n", callback->callback, callback->userdata);
+        callback->callback(0, callback->userdata);
+        TRACE("handler %p returned\n", callback->callback);
+
+        ptr = list_next( &tls_callbacks, ptr );
+        if (!--callback->count)  /* removed during execution */
+        {
+            list_remove( &callback->entry );
+            HeapFree( GetProcessHeap(), 0, callback );
+        }
+    }
+    LeaveCriticalSection( &tls_callback_section );
+}
+
+static const CUuuid UUID_Unknown1                   = {{0x6B, 0xD5, 0xFB, 0x6C, 0x5B, 0xF4, 0xE7, 0x4A,
+                                                        0x89, 0x87, 0xD9, 0x39, 0x12, 0xFD, 0x9D, 0xF9}};
+static const CUuuid UUID_Unknown2                   = {{0xA0, 0x94, 0x79, 0x8C, 0x2E, 0x74, 0x2E, 0x74,
+                                                        0x93, 0xF2, 0x08, 0x00, 0x20, 0x0C, 0x0A, 0x66}};
+static const CUuuid UUID_Unknown3                   = {{0x42, 0xD8, 0x5A, 0x81, 0x23, 0xF6, 0xCB, 0x47,
+                                                        0x82, 0x98, 0xF6, 0xE7, 0x8A, 0x3A, 0xEC, 0xDC}};
+static const CUuuid UUID_Unknown4                   = {{0xC6, 0x93, 0x33, 0x6E, 0x11, 0x21, 0xDF, 0x11,
+                                                        0xA8, 0xC3, 0x68, 0xF3, 0x55, 0xD8, 0x95, 0x93}};
+static const CUuuid UUID_Unknown5                   = {{0x0C, 0xA5, 0x0B, 0x8C, 0x10, 0x04, 0x92, 0x9A,
+                                                        0x89, 0xA7, 0xD0, 0xDF, 0x10, 0xE7, 0x72, 0x86}};
+static const CUuuid UUID_TlsNotifyInterface         = {{0x19, 0x5B, 0xCB, 0xF4, 0xD6, 0x7D, 0x02, 0x4A,
+                                                        0xAC, 0xC5, 0x1D, 0x29, 0xCE, 0xA6, 0x31, 0xAE}};
+
+struct cuda_table
+{
+    int size;
+    void *functions[0];
+};
+
+/*
+ * Unknown1
+ */
+struct Unknown1_table
+{
+    int size;
+    void* (WINAPI *func0)(void *param0, void *param1);
+    void* (WINAPI *func1)(void *param0, void *param1);
+    void* (WINAPI *func2)(void *param0, void *param1);
+    void* (WINAPI *func3)(void *param0, void *param1);
+    void* (WINAPI *func4)(void *param0);
+};
+static const struct
+{
+    int size;
+    void* (*func0)(void *param0, void *param1);
+    void* (*func1)(void *param0, void *param1);
+    void* (*func2)(void *param0, void *param1);
+    void* (*func3)(void *param0, void *param1);
+    void* (*func4)(void *param0);
+} *Unknown1_orig = NULL;
+
+/*
+ * Unknown2
+ */
+struct Unknown2_table
+{
+    int size;
+    void* (WINAPI *func0)(void *param0, void *param1);
+    void* (WINAPI *func1)(void *param0, void *param1);
+    void* (WINAPI *func2)(void *param0, void *param1, void *param2);
+    void* (WINAPI *func3)(void *param0, void *param1);
+    void* (WINAPI *func4)(void *param0, void *param1);
+    void* (WINAPI *func5)(void *param0, void *param1);
+};
+static const struct
+{
+    int size;
+    void* (*func0)(void *param0, void *param1);
+    void* (*func1)(void *param0, void *param1);
+    void* (*func2)(void *param0, void *param1, void *param2);
+    void* (*func3)(void *param0, void *param1);
+    void* (*func4)(void *param0, void *param1);
+    void* (*func5)(void *param0, void *param1);
+} *Unknown2_orig = NULL;
+
+/*
+ * Unknown3
+ */
+struct Unknown3_table
+{
+    int size;
+    void* (WINAPI *func0)(void *param0);
+    void* (WINAPI *func1)(void *param0);
+};
+static const struct
+{
+    int size;
+    void* (*func0)(void *param0);
+    void* (*func1)(void *param0);
+} *Unknown3_orig = NULL;
+
+/*
+ * Unknown4
+ */
+struct Unknown4_table
+{
+    void* (WINAPI *func0)(void *param0, void *param1, void *param2, void *param3);
+    void* (WINAPI *func1)(void *param0, void *param1);
+    void* (WINAPI *func2)(void *param0, void *param1, void *param2);
+};
+static const struct
+{
+    void* (*func0)(void *param0, void *param1, void *param2, void *param3);
+    void* (*func1)(void *param0, void *param1);
+    void* (*func2)(void *param0, void *param1, void *param2);
+} *Unknown4_orig = NULL;
+
+/*
+ * TlsNotifyInterface
+ */
+struct TlsNotifyInterface_table
+{
+    int size;
+    CUresult (WINAPI *Set)(void **handle, void *callback, void *data);
+    CUresult (WINAPI *Remove)(void *handle, void *param1);
+};
+
+/*
+ * Unknown5
+ */
+struct Unknown5_table
+{
+    int size;
+    void* (WINAPI *func0)(void *param0, void *param1, void *param2);
+};
+static const struct
+{
+    int size;
+    void* (*func0)(void *param0, void *param1, void *param2);
+} *Unknown5_orig = NULL;
+
+
+static void* WINAPI Unknown1_func0_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown1_orig->func0(param0, param1);
+}
+
+static void* WINAPI Unknown1_func1_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown1_orig->func1(param0, param1);
+}
+
+static void* WINAPI Unknown1_func2_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown1_orig->func2(param0, param1);
+}
+
+static void* WINAPI Unknown1_func3_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown1_orig->func3(param0, param1);
+}
+
+static void* WINAPI Unknown1_func4_relay(void *param0)
+{
+    TRACE("(%p)\n", param0);
+    return Unknown1_orig->func4(param0);
+}
+
+struct Unknown1_table Unknown1_Impl =
+{
+    sizeof(struct Unknown1_table),
+    Unknown1_func0_relay,
+    Unknown1_func1_relay,
+    Unknown1_func2_relay,
+    Unknown1_func3_relay,
+    Unknown1_func4_relay,
+};
+
+static void* WINAPI Unknown2_func0_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown2_orig->func0(param0, param1);
+}
+
+static void* WINAPI Unknown2_func1_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown2_orig->func1(param0, param1);
+}
+
+static void* WINAPI Unknown2_func2_relay(void *param0, void *param1, void *param2)
+{
+    TRACE("(%p, %p, %p)\n", param0, param1, param2);
+    return Unknown2_orig->func2(param0, param1, param2);
+}
+
+static void* WINAPI Unknown2_func3_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown2_orig->func3(param0, param1);
+}
+
+static void* WINAPI Unknown2_func4_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown2_orig->func4(param0, param1);
+}
+
+static void* WINAPI Unknown2_func5_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown2_orig->func5(param0, param1);
+}
+
+struct Unknown2_table Unknown2_Impl =
+{
+    sizeof(struct Unknown2_table),
+    Unknown2_func0_relay,
+    Unknown2_func1_relay,
+    Unknown2_func2_relay,
+    Unknown2_func3_relay,
+    Unknown2_func4_relay,
+    Unknown2_func5_relay,
+};
+
+static void* WINAPI Unknown3_func0_relay(void *param0)
+{
+    TRACE("(%p)\n", param0);
+    return Unknown3_orig->func0(param0);
+}
+
+static void* WINAPI Unknown3_func1_relay(void *param0)
+{
+    TRACE("(%p)\n", param0);
+    return Unknown3_orig->func1(param0);
+}
+
+static struct Unknown3_table Unknown3_Impl =
+{
+    sizeof(struct Unknown3_table),
+    Unknown3_func0_relay,
+    Unknown3_func1_relay,
+};
+
+static void* WINAPI Unknown4_func0_relay(void *param0, void *param1, void *param2, void *param3)
+{
+    TRACE("(%p, %p, %p, %p)\n", param0, param1, param2, param3);
+    return Unknown4_orig->func0(param0, param1, param2, param3);
+}
+
+static void* WINAPI Unknown4_func1_relay(void *param0, void *param1)
+{
+    TRACE("(%p, %p)\n", param0, param1);
+    return Unknown4_orig->func1(param0, param1);
+}
+
+static void* WINAPI Unknown4_func2_relay(void *param0, void *param1, void *param2)
+{
+    TRACE("(%p, %p, %p)\n", param0, param1, param2);
+    return Unknown4_orig->func2(param0, param1, param2);
+}
+
+struct Unknown4_table Unknown4_Impl =
+{
+    Unknown4_func0_relay,
+    Unknown4_func1_relay,
+    Unknown4_func2_relay,
+};
+
+static void* WINAPI Unknown5_func0_relay(void *param0, void *param1, void *param2)
+{
+    TRACE("(%p, %p, %p)\n", param0, param1, param2);
+    return Unknown5_orig->func0(param0, param1, param2);
+}
+
+struct Unknown5_table Unknown5_Impl =
+{
+    sizeof(struct Unknown5_table),
+    Unknown5_func0_relay,
+};
+
+static CUresult WINAPI TlsNotifyInterface_Set(void **handle, void *callback, void *userdata)
+{
+    struct tls_callback_entry *new_entry;
+
+    TRACE("(%p, %p, %p)\n", handle, callback, userdata);
+
+    new_entry = HeapAlloc( GetProcessHeap(), 0, sizeof(*new_entry) );
+    if (!new_entry)
+        return CUDA_ERROR_OUT_OF_MEMORY;
+
+    new_entry->callback = callback;
+    new_entry->userdata = userdata;
+    new_entry->count = 1;
+
+    EnterCriticalSection( &tls_callback_section );
+    list_add_tail( &tls_callbacks, &new_entry->entry );
+    LeaveCriticalSection( &tls_callback_section );
+
+    *handle = new_entry;
+    return CUDA_SUCCESS;
+}
+
+static CUresult WINAPI TlsNotifyInterface_Remove(void *handle, void *param1)
+{
+    CUresult ret = CUDA_ERROR_INVALID_VALUE;
+    struct tls_callback_entry *to_free = NULL;
+    struct list *ptr;
+
+    TRACE("(%p, %p)\n", handle, param1);
+
+    if (param1)
+        FIXME("semi stub: param1 != 0 not supported.\n");
+
+    EnterCriticalSection( &tls_callback_section );
+    LIST_FOR_EACH( ptr, &tls_callbacks )
+    {
+        struct tls_callback_entry *callback = LIST_ENTRY( ptr, struct tls_callback_entry, entry );
+        if (callback == handle)
+        {
+            if (!--callback->count)
+            {
+                list_remove( ptr );
+                to_free = callback;
+            }
+            ret = CUDA_SUCCESS;
+            break;
+        }
+    }
+    LeaveCriticalSection( &tls_callback_section );
+    HeapFree( GetProcessHeap(), 0, to_free );
+    return ret;
+}
+
+struct TlsNotifyInterface_table TlsNotifyInterface_Impl =
+{
+    sizeof(struct TlsNotifyInterface_table),
+    TlsNotifyInterface_Set,
+    TlsNotifyInterface_Remove,
+};
+
+static BOOL cuda_check_table(const struct cuda_table *orig, struct cuda_table *impl, const char *name)
+{
+    if (!orig)
+        return FALSE;
+
+    /* FIXME: better check for size, verify that function pointers are != NULL */
+
+    if (orig->size > impl->size)
+    {
+        FIXME("WARNING: Your CUDA version supports a newer interface for %s then the Wine implementation.\n", name);
+    }
+    else if (orig->size < impl->size)
+    {
+        FIXME("Your CUDA version supports only an older interface for %s, downgrading version.\n", name);
+        impl->size = orig->size;
+    }
+
+    return TRUE;
+}
+
+static inline BOOL cuda_equal_uuid(const CUuuid *id1, const CUuuid *id2)
+{
+    return !memcmp(id1, id2, sizeof(CUuuid));
+}
+
+static char* cuda_print_uuid(const CUuuid *id, char *buffer, int size)
+{
+    snprintf(buffer, size, "{0x%02X, 0x%02X, 0x%02X, 0x%02X, 0x%02X, 0x%02X, 0x%02X, 0x%02X, " \
+                            "0x%02X, 0x%02X, 0x%02X, 0x%02X, 0x%02X, 0x%02X, 0x%02X, 0x%02X}",
+             id->bytes[0] & 0xFF, id->bytes[1] & 0xFF, id->bytes[2] & 0xFF, id->bytes[3] & 0xFF,
+             id->bytes[4] & 0xFF, id->bytes[5] & 0xFF, id->bytes[6] & 0xFF, id->bytes[7] & 0xFF,
+             id->bytes[8] & 0xFF, id->bytes[9] & 0xFF, id->bytes[10] & 0xFF, id->bytes[11] & 0xFF,
+             id->bytes[12] & 0xFF, id->bytes[13] & 0xFF, id->bytes[14] & 0xFF, id->bytes[15] & 0xFF);
+    return buffer;
+}
+
+CUresult cuda_get_table(const void **table, const CUuuid *uuid, const void *orig_table, CUresult orig_result)
+{
+    char buffer[128];
+
+    if (cuda_equal_uuid(uuid, &UUID_Unknown1))
+    {
+        if (orig_result)
+            return orig_result;
+        if (!cuda_check_table(orig_table, (void *)&Unknown1_Impl, "Unknown1"))
+            return CUDA_ERROR_UNKNOWN;
+
+        Unknown1_orig = orig_table;
+        *table = (void *)&Unknown1_Impl;
+        return CUDA_SUCCESS;
+    }
+    else if (cuda_equal_uuid(uuid, &UUID_Unknown2))
+    {
+        if (orig_result)
+            return orig_result;
+        if (!cuda_check_table(orig_table, (void *)&Unknown2_Impl, "Unknown2"))
+            return CUDA_ERROR_UNKNOWN;
+
+        Unknown2_orig = orig_table;
+        *table = (void *)&Unknown2_Impl;
+        return CUDA_SUCCESS;
+    }
+    else if (cuda_equal_uuid(uuid, &UUID_Unknown3))
+    {
+        if (orig_result)
+            return orig_result;
+        if (!cuda_check_table(orig_table, (void *)&Unknown3_Impl, "Unknown3"))
+            return CUDA_ERROR_UNKNOWN;
+
+        Unknown3_orig = orig_table;
+        *table = (void *)&Unknown3_Impl;
+        return CUDA_SUCCESS;
+    }
+    else if (cuda_equal_uuid(uuid, &UUID_Unknown4))
+    {
+        if (orig_result)
+            return orig_result;
+        if (!orig_table)
+            return CUDA_ERROR_UNKNOWN;
+
+        Unknown4_orig = orig_table;
+        *table = (void *)&Unknown4_Impl;
+        return CUDA_SUCCESS;
+    }
+    else if (cuda_equal_uuid(uuid, &UUID_Unknown5))
+    {
+        if (orig_result)
+            return orig_result;
+        if (!cuda_check_table(orig_table, (void *)&Unknown5_Impl, "Unknown5"))
+            return CUDA_ERROR_UNKNOWN;
+
+        Unknown5_orig = orig_table;
+        *table = (void *)&Unknown5_Impl;
+        return CUDA_SUCCESS;
+    }
+    else if (cuda_equal_uuid(uuid, &UUID_TlsNotifyInterface))
+    {
+        /* the following interface is not implemented in the Linux
+         * CUDA driver, we provide a replacement implementation */
+        *table = (void *)&TlsNotifyInterface_Impl;
+        return CUDA_SUCCESS;
+    }
+
+    FIXME("Unknown UUID: %s, error: %d\n", cuda_print_uuid(uuid, buffer, sizeof(buffer)), orig_result);
+    return CUDA_ERROR_UNKNOWN;
+}
diff --git a/dlls/nvcuda/nvcuda.c b/dlls/nvcuda/nvcuda.c
index 8b7eb9c..3ef7c35 100644
--- a/dlls/nvcuda/nvcuda.c
+++ b/dlls/nvcuda/nvcuda.c
@@ -1,4 +1,5 @@
 /*
+ * Copyright (C) 2014-2015 Michael Müller
  * Copyright (C) 2014-2015 Sebastian Lackner
  *
  * This library is free software; you can redistribute it and/or
@@ -17,19 +18,2105 @@
  */
 
 #include "config.h"
+#include "wine/port.h"
+
 #include <stdarg.h>
 
-#include "windef.h"
-#include "winbase.h"
+#include "windef.h"
+#include "winbase.h"
+#include "wine/library.h"
+#include "wine/debug.h"
+#include "wine/wgl.h"
+#include "cuda.h"
+#include "nvcuda.h"
+
+#if defined(__x86_64) || defined(AMD64) || defined(_M_AMD64)
+#define DEV_PTR "%llu"
+#else
+#define DEV_PTR "%u"
+#endif
+
+WINE_DEFAULT_DEBUG_CHANNEL(nvcuda);
+
+static CUresult (*pcuArray3DCreate)(CUarray *pHandle, const CUDA_ARRAY3D_DESCRIPTOR *pAllocateArray);
+static CUresult (*pcuArray3DCreate_v2)(CUarray *pHandle, const CUDA_ARRAY3D_DESCRIPTOR *pAllocateArray);
+static CUresult (*pcuArray3DGetDescriptor)(CUDA_ARRAY3D_DESCRIPTOR *pArrayDescriptor, CUarray hArray);
+static CUresult (*pcuArray3DGetDescriptor_v2)(CUDA_ARRAY3D_DESCRIPTOR *pArrayDescriptor, CUarray hArray);
+static CUresult (*pcuArrayCreate)(CUarray *pHandle, const CUDA_ARRAY_DESCRIPTOR *pAllocateArray);
+static CUresult (*pcuArrayCreate_v2)(CUarray *pHandle, const CUDA_ARRAY_DESCRIPTOR *pAllocateArray);
+static CUresult (*pcuArrayDestroy)(CUarray hArray);
+static CUresult (*pcuArrayGetDescriptor)(CUDA_ARRAY_DESCRIPTOR *pArrayDescriptor, CUarray hArray);
+static CUresult (*pcuArrayGetDescriptor_v2)(CUDA_ARRAY_DESCRIPTOR *pArrayDescriptor, CUarray hArray);
+static CUresult (*pcuCtxAttach)(CUcontext *pctx, unsigned int flags);
+static CUresult (*pcuCtxCreate)(CUcontext *pctx, unsigned int flags, CUdevice dev);
+static CUresult (*pcuCtxCreate_v2)(CUcontext *pctx, unsigned int flags, CUdevice dev);
+static CUresult (*pcuCtxDestroy)(CUcontext ctx);
+static CUresult (*pcuCtxDestroy_v2)(CUcontext ctx);
+static CUresult (*pcuCtxDetach)(CUcontext ctx);
+static CUresult (*pcuCtxDisablePeerAccess)(CUcontext peerContext);
+static CUresult (*pcuCtxEnablePeerAccess)(CUcontext peerContext, unsigned int Flags);
+static CUresult (*pcuCtxGetApiVersion)(CUcontext ctx, unsigned int *version);
+static CUresult (*pcuCtxGetCacheConfig)(CUfunc_cache *pconfig);
+static CUresult (*pcuCtxGetCurrent)(CUcontext *pctx);
+static CUresult (*pcuCtxGetDevice)(CUdevice *device);
+static CUresult (*pcuCtxGetLimit)(size_t *pvalue, CUlimit limit);
+static CUresult (*pcuCtxGetSharedMemConfig)(CUsharedconfig *pConfig);
+static CUresult (*pcuCtxGetStreamPriorityRange)(int *leastPriority, int *greatestPriority);
+static CUresult (*pcuCtxPopCurrent)(CUcontext *pctx);
+static CUresult (*pcuCtxPopCurrent_v2)(CUcontext *pctx);
+static CUresult (*pcuCtxPushCurrent)(CUcontext ctx);
+static CUresult (*pcuCtxPushCurrent_v2)(CUcontext ctx);
+static CUresult (*pcuCtxSetCacheConfig)(CUfunc_cache config);
+static CUresult (*pcuCtxSetCurrent)(CUcontext ctx);
+static CUresult (*pcuCtxSetLimit)(CUlimit limit, size_t value);
+static CUresult (*pcuCtxSetSharedMemConfig)(CUsharedconfig config);
+static CUresult (*pcuCtxSynchronize)(void);
+static CUresult (*pcuDeviceCanAccessPeer)(int *canAccessPeer, CUdevice dev, CUdevice peerDev);
+static CUresult (*pcuDeviceComputeCapability)(int *major, int *minor, CUdevice dev);
+static CUresult (*pcuDeviceGet)(CUdevice *device, int ordinal);
+static CUresult (*pcuDeviceGetAttribute)(int *pi, CUdevice_attribute attrib, CUdevice dev);
+static CUresult (*pcuDeviceGetByPCIBusId)(CUdevice *dev, const char *pciBusId);
+static CUresult (*pcuDeviceGetCount)(int *count);
+static CUresult (*pcuDeviceGetName)(char *name, int len, CUdevice dev);
+static CUresult (*pcuDeviceGetPCIBusId)(char *pciBusId, int len, CUdevice dev);
+static CUresult (*pcuDeviceGetProperties)(CUdevprop *prop, CUdevice dev);
+static CUresult (*pcuDeviceTotalMem)(size_t *bytes, CUdevice dev);
+static CUresult (*pcuDeviceTotalMem_v2)(size_t *bytes, CUdevice dev);
+static CUresult (*pcuDriverGetVersion)(int *);
+static CUresult (*pcuEventCreate)(CUevent *phEvent, unsigned int Flags);
+static CUresult (*pcuEventDestroy)(CUevent hEvent);
+static CUresult (*pcuEventDestroy_v2)(CUevent hEvent);
+static CUresult (*pcuEventElapsedTime)(float *pMilliseconds, CUevent hStart, CUevent hEnd);
+static CUresult (*pcuEventQuery)(CUevent hEvent);
+static CUresult (*pcuEventRecord)(CUevent hEvent, CUstream hStream);
+static CUresult (*pcuEventSynchronize)(CUevent hEvent);
+static CUresult (*pcuFuncGetAttribute)(int *pi, CUfunction_attribute attrib, CUfunction hfunc);
+static CUresult (*pcuFuncSetBlockShape)(CUfunction hfunc, int x, int y, int z);
+static CUresult (*pcuFuncSetCacheConfig)(CUfunction hfunc, CUfunc_cache config);
+static CUresult (*pcuFuncSetSharedMemConfig)(CUfunction hfunc, CUsharedconfig config);
+static CUresult (*pcuFuncSetSharedSize)(CUfunction hfunc, unsigned int bytes);
+static CUresult (*pcuGLCtxCreate)(CUcontext *pCtx, unsigned int Flags, CUdevice device);
+static CUresult (*pcuGLCtxCreate_v2)(CUcontext *pCtx, unsigned int Flags, CUdevice device);
+static CUresult (*pcuGLGetDevices)(unsigned int *pCudaDeviceCount, CUdevice *pCudaDevices,
+                                   unsigned int cudaDeviceCount, CUGLDeviceList deviceList);
+static CUresult (*pcuGLInit)(void);
+static CUresult (*pcuGLMapBufferObject)(CUdeviceptr *dptr, size_t *size, GLuint buffer);
+static CUresult (*pcuGLMapBufferObjectAsync)(CUdeviceptr *dptr, size_t *size, GLuint buffer, CUstream hStream);
+static CUresult (*pcuGLMapBufferObjectAsync_v2)(CUdeviceptr *dptr, size_t *size, GLuint buffer, CUstream hStream);
+static CUresult (*pcuGLMapBufferObject_v2)(CUdeviceptr *dptr, size_t *size, GLuint buffer);
+static CUresult (*pcuGLRegisterBufferObject)(GLuint buffer);
+static CUresult (*pcuGLSetBufferObjectMapFlags)(GLuint buffer, unsigned int Flags);
+static CUresult (*pcuGLUnmapBufferObject)(GLuint buffer);
+static CUresult (*pcuGLUnmapBufferObjectAsync)(GLuint buffer, CUstream hStream);
+static CUresult (*pcuGLUnregisterBufferObject)(GLuint buffer);
+static CUresult (*pcuGetErrorName)(CUresult error, const char **pStr);
+static CUresult (*pcuGetErrorString)(CUresult error, const char **pStr);
+static CUresult (*pcuGetExportTable)(const void**, const CUuuid*);
+static CUresult (*pcuGraphicsGLRegisterBuffer)(CUgraphicsResource *pCudaResource, GLuint buffer, unsigned int Flags);
+static CUresult (*pcuGraphicsGLRegisterImage)(CUgraphicsResource *pCudaResource, GLuint image, GLenum target, unsigned int Flags);
+static CUresult (*pcuGraphicsMapResources)(unsigned int count, CUgraphicsResource *resources, CUstream hStream);
+static CUresult (*pcuGraphicsResourceGetMappedMipmappedArray)(CUmipmappedArray *pMipmappedArray, CUgraphicsResource resource);
+static CUresult (*pcuGraphicsResourceGetMappedPointer)(CUdeviceptr *pDevPtr, size_t *pSize, CUgraphicsResource resource);
+static CUresult (*pcuGraphicsResourceGetMappedPointer_v2)(CUdeviceptr *pDevPtr, size_t *pSize, CUgraphicsResource resource);
+static CUresult (*pcuGraphicsResourceSetMapFlags)(CUgraphicsResource resource, unsigned int flags);
+static CUresult (*pcuGraphicsSubResourceGetMappedArray)(CUarray *pArray, CUgraphicsResource resource,
+                                                        unsigned int arrayIndex, unsigned int mipLevel);
+static CUresult (*pcuGraphicsUnmapResources)(unsigned int count, CUgraphicsResource *resources, CUstream hStream);
+static CUresult (*pcuGraphicsUnregisterResource)(CUgraphicsResource resource);
+static CUresult (*pcuInit)(unsigned int);
+static CUresult (*pcuIpcCloseMemHandle)(CUdeviceptr dptr);
+static CUresult (*pcuIpcGetEventHandle)(CUipcEventHandle *pHandle, CUevent event);
+static CUresult (*pcuIpcGetMemHandle)(CUipcMemHandle *pHandle, CUdeviceptr dptr);
+static CUresult (*pcuIpcOpenEventHandle)(CUevent *phEvent, CUipcEventHandle handle);
+static CUresult (*pcuIpcOpenMemHandle)(CUdeviceptr *pdptr, CUipcMemHandle handle, unsigned int Flags);
+static CUresult (*pcuLaunch)(CUfunction f);
+static CUresult (*pcuLaunchGrid)(CUfunction f, int grid_width, int grid_height);
+static CUresult (*pcuLaunchGridAsync)(CUfunction f, int grid_width, int grid_height, CUstream hStream);
+static CUresult (*pcuLaunchKernel)(CUfunction f, unsigned int gridDimX, unsigned int gridDimY, unsigned int gridDimZ,
+                                   unsigned int blockDimX, unsigned int blockDimY, unsigned int blockDimZ,
+                                   unsigned int sharedMemBytes, CUstream hStream, void **kernelParams, void **extra);
+static CUresult (*pcuLinkAddData)(CUlinkState state, CUjitInputType type, void *data, size_t size, const char *name,
+                                  unsigned int numOptions, CUjit_option *options, void **optionValues);
+static CUresult (*pcuLinkComplete)(CUlinkState state, void **cubinOut, size_t *sizeOut);
+static CUresult (*pcuLinkCreate)(unsigned int numOptions, CUjit_option *options, void **optionValues, CUlinkState *stateOut);
+static CUresult (*pcuLinkDestroy)(CUlinkState state);
+static CUresult (*pcuMemAlloc)(CUdeviceptr *dptr, unsigned int bytesize);
+static CUresult (*pcuMemAllocHost)(void **pp, size_t bytesize);
+static CUresult (*pcuMemAllocHost_v2)(void **pp, size_t bytesize);
+static CUresult (*pcuMemAllocManaged)(CUdeviceptr *dptr, size_t bytesize, unsigned int flags);
+static CUresult (*pcuMemAllocPitch)(CUdeviceptr *dptr, size_t *pPitch, size_t WidthInBytes, size_t Height, unsigned int ElementSizeBytes);
+static CUresult (*pcuMemAllocPitch_v2)(CUdeviceptr *dptr, size_t *pPitch, size_t WidthInBytes, size_t Height, unsigned int ElementSizeBytes);
+static CUresult (*pcuMemAlloc_v2)(CUdeviceptr *dptr, unsigned int bytesize);
+static CUresult (*pcuMemFree)(CUdeviceptr dptr);
+static CUresult (*pcuMemFreeHost)(void *p);
+static CUresult (*pcuMemFree_v2)(CUdeviceptr dptr);
+static CUresult (*pcuMemGetAddressRange)(CUdeviceptr *pbase, unsigned int *psize, CUdeviceptr dptr);
+static CUresult (*pcuMemGetAddressRange_v2)(CUdeviceptr *pbase, unsigned int *psize, CUdeviceptr dptr);
+static CUresult (*pcuMemGetInfo)(size_t *free, size_t *total);
+static CUresult (*pcuMemGetInfo_v2)(size_t *free, size_t *total);
+static CUresult (*pcuMemHostAlloc)(void **pp, size_t bytesize, unsigned int Flags);
+static CUresult (*pcuMemHostGetDevicePointer)(CUdeviceptr *pdptr, void *p, unsigned int Flags);
+static CUresult (*pcuMemHostGetDevicePointer_v2)(CUdeviceptr *pdptr, void *p, unsigned int Flags);
+static CUresult (*pcuMemHostGetFlags)(unsigned int *pFlags, void *p);
+static CUresult (*pcuMemHostRegister)(void *p, size_t bytesize, unsigned int Flags);
+static CUresult (*pcuMemHostUnregister)(void *p);
+static CUresult (*pcuMemcpy)(CUdeviceptr dst, CUdeviceptr src, size_t ByteCount);
+static CUresult (*pcuMemcpy2D)(const CUDA_MEMCPY2D *pCopy);
+static CUresult (*pcuMemcpy2DAsync)(const CUDA_MEMCPY2D *pCopy, CUstream hStream);
+static CUresult (*pcuMemcpy2DAsync_v2)(const CUDA_MEMCPY2D *pCopy, CUstream hStream);
+static CUresult (*pcuMemcpy2DUnaligned)(const CUDA_MEMCPY2D *pCopy);
+static CUresult (*pcuMemcpy2DUnaligned_v2)(const CUDA_MEMCPY2D *pCopy);
+static CUresult (*pcuMemcpy2D_v2)(const CUDA_MEMCPY2D *pCopy);
+static CUresult (*pcuMemcpy3D)(const CUDA_MEMCPY3D *pCopy);
+static CUresult (*pcuMemcpy3DAsync)(const CUDA_MEMCPY3D *pCopy, CUstream hStream);
+static CUresult (*pcuMemcpy3DAsync_v2)(const CUDA_MEMCPY3D *pCopy, CUstream hStream);
+static CUresult (*pcuMemcpy3DPeer)(const CUDA_MEMCPY3D_PEER *pCopy);
+static CUresult (*pcuMemcpy3DPeerAsync)(const CUDA_MEMCPY3D_PEER *pCopy, CUstream hStream);
+static CUresult (*pcuMemcpy3D_v2)(const CUDA_MEMCPY3D *pCopy);
+static CUresult (*pcuMemcpyAsync)(CUdeviceptr dst, CUdeviceptr src, size_t ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyAtoA)(CUarray dstArray, size_t dstOffset, CUarray srcArray, size_t srcOffset, size_t ByteCount);
+static CUresult (*pcuMemcpyAtoA_v2)(CUarray dstArray, size_t dstOffset, CUarray srcArray, size_t srcOffset, size_t ByteCount);
+static CUresult (*pcuMemcpyAtoD)(CUdeviceptr dstDevice, CUarray srcArray, size_t srcOffset, size_t ByteCount);
+static CUresult (*pcuMemcpyAtoD_v2)(CUdeviceptr dstDevice, CUarray srcArray, size_t srcOffset, size_t ByteCount);
+static CUresult (*pcuMemcpyAtoH)(void *dstHost, CUarray srcArray, size_t srcOffset, size_t ByteCount);
+static CUresult (*pcuMemcpyAtoHAsync)(void *dstHost, CUarray srcArray, size_t srcOffset, size_t ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyAtoHAsync_v2)(void *dstHost, CUarray srcArray, size_t srcOffset, size_t ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyAtoH_v2)(void *dstHost, CUarray srcArray, size_t srcOffset, size_t ByteCount);
+static CUresult (*pcuMemcpyDtoA)(CUarray dstArray, size_t dstOffset, CUdeviceptr srcDevice, size_t ByteCount);
+static CUresult (*pcuMemcpyDtoA_v2)(CUarray dstArray, size_t dstOffset, CUdeviceptr srcDevice, size_t ByteCount);
+static CUresult (*pcuMemcpyDtoD)(CUdeviceptr dstDevice, CUdeviceptr srcDevice, unsigned int ByteCount);
+static CUresult (*pcuMemcpyDtoDAsync)(CUdeviceptr dstDevice, CUdeviceptr srcDevice, unsigned int ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyDtoDAsync_v2)(CUdeviceptr dstDevice, CUdeviceptr srcDevice, unsigned int ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyDtoD_v2)(CUdeviceptr dstDevice, CUdeviceptr srcDevice, unsigned int ByteCount);
+static CUresult (*pcuMemcpyDtoH)(void *dstHost, CUdeviceptr srcDevice, unsigned int ByteCount);
+static CUresult (*pcuMemcpyDtoHAsync)(void *dstHost, CUdeviceptr srcDevice, unsigned int ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyDtoHAsync_v2)(void *dstHost, CUdeviceptr srcDevice, unsigned int ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyDtoH_v2)(void *dstHost, CUdeviceptr srcDevice, unsigned int ByteCount);
+static CUresult (*pcuMemcpyHtoA)(CUarray dstArray, size_t dstOffset, const void *srcHost, size_t ByteCount);
+static CUresult (*pcuMemcpyHtoAAsync)(CUarray dstArray, size_t dstOffset, const void *srcHost, size_t ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyHtoAAsync_v2)(CUarray dstArray, size_t dstOffset, const void *srcHost, size_t ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyHtoA_v2)(CUarray dstArray, size_t dstOffset, const void *srcHost, size_t ByteCount);
+static CUresult (*pcuMemcpyHtoD)(CUdeviceptr dstDevice, const void *srcHost, size_t ByteCount);
+static CUresult (*pcuMemcpyHtoDAsync)(CUdeviceptr dstDevice, const void *srcHost, size_t ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyHtoDAsync_v2)(CUdeviceptr dstDevice, const void *srcHost, size_t ByteCount, CUstream hStream);
+static CUresult (*pcuMemcpyHtoD_v2)(CUdeviceptr dstDevice, const void *srcHost, size_t ByteCount);
+static CUresult (*pcuMemcpyPeer)(CUdeviceptr dstDevice, CUcontext dstContext, CUdeviceptr srcDevice, CUcontext srcContext, size_t ByteCount);
+static CUresult (*pcuMemcpyPeerAsync)(CUdeviceptr dstDevice, CUcontext dstContext, CUdeviceptr srcDevice,
+                                      CUcontext srcContext, size_t ByteCount, CUstream hStream);
+static CUresult (*pcuMemsetD16)(CUdeviceptr dstDevice, unsigned short us, size_t N);
+static CUresult (*pcuMemsetD16Async)(CUdeviceptr dstDevice, unsigned short us, size_t N, CUstream hStream);
+static CUresult (*pcuMemsetD16_v2)(CUdeviceptr dstDevice, unsigned short us, size_t N);
+static CUresult (*pcuMemsetD2D16)(CUdeviceptr dstDevice, size_t dstPitch, unsigned short us, size_t Width, size_t Height);
+static CUresult (*pcuMemsetD2D16Async)(CUdeviceptr dstDevice, size_t dstPitch, unsigned short us, size_t Width, size_t Height, CUstream hStream);
+static CUresult (*pcuMemsetD2D16_v2)(CUdeviceptr dstDevice, size_t dstPitch, unsigned short us, size_t Width, size_t Height);
+static CUresult (*pcuMemsetD2D32)(CUdeviceptr dstDevice, size_t dstPitch, unsigned int ui, size_t Width, size_t Height);
+static CUresult (*pcuMemsetD2D32Async)(CUdeviceptr dstDevice, size_t dstPitch, unsigned int ui, size_t Width, size_t Height, CUstream hStream);
+static CUresult (*pcuMemsetD2D32_v2)(CUdeviceptr dstDevice, size_t dstPitch, unsigned int ui, size_t Width, size_t Height);
+static CUresult (*pcuMemsetD2D8)(CUdeviceptr dstDevice, unsigned int dstPitch, unsigned char uc, unsigned int Width, unsigned int Height);
+static CUresult (*pcuMemsetD2D8Async)(CUdeviceptr dstDevice, size_t dstPitch, unsigned char uc, size_t Width, size_t Height, CUstream hStream);
+static CUresult (*pcuMemsetD2D8_v2)(CUdeviceptr dstDevice, unsigned int dstPitch, unsigned char uc, unsigned int Width, unsigned int Height);
+static CUresult (*pcuMemsetD32)(CUdeviceptr dstDevice, unsigned int ui, size_t N);
+static CUresult (*pcuMemsetD32Async)(CUdeviceptr dstDevice, unsigned int ui, size_t N, CUstream hStream);
+static CUresult (*pcuMemsetD32_v2)(CUdeviceptr dstDevice, unsigned int ui, size_t N);
+static CUresult (*pcuMemsetD8)(CUdeviceptr dstDevice, unsigned char uc, unsigned int N);
+static CUresult (*pcuMemsetD8Async)(CUdeviceptr dstDevice, unsigned char uc, size_t N, CUstream hStream);
+static CUresult (*pcuMemsetD8_v2)(CUdeviceptr dstDevice, unsigned char uc, unsigned int N);
+static CUresult (*pcuMipmappedArrayCreate)(CUmipmappedArray *pHandle, const CUDA_ARRAY3D_DESCRIPTOR *pMipmappedArrayDesc,
+                                           unsigned int numMipmapLevels);
+static CUresult (*pcuMipmappedArrayDestroy)(CUmipmappedArray hMipmappedArray);
+static CUresult (*pcuMipmappedArrayGetLevel)(CUarray *pLevelArray, CUmipmappedArray hMipmappedArray, unsigned int level);
+static CUresult (*pcuModuleGetFunction)(CUfunction *hfunc, CUmodule hmod, const char *name);
+static CUresult (*pcuModuleGetGlobal)(CUdeviceptr *dptr, size_t *bytes, CUmodule hmod, const char *name);
+static CUresult (*pcuModuleGetGlobal_v2)(CUdeviceptr *dptr, size_t *bytes, CUmodule hmod, const char *name);
+static CUresult (*pcuModuleGetSurfRef)(CUsurfref *pSurfRef, CUmodule hmod, const char *name);
+static CUresult (*pcuModuleGetTexRef)(CUtexref *pTexRef, CUmodule hmod, const char *name);
+static CUresult (*pcuModuleLoadData)(CUmodule *module, const void *image);
+static CUresult (*pcuModuleLoadDataEx)(CUmodule *module, const void *image, unsigned int numOptions, CUjit_option *options, void **optionValues);
+static CUresult (*pcuModuleLoadFatBinary)(CUmodule *module, const void *fatCubin);
+static CUresult (*pcuModuleUnload)(CUmodule hmod);
+static CUresult (*pcuParamSetSize)(CUfunction hfunc, unsigned int numbytes);
+static CUresult (*pcuParamSetTexRef)(CUfunction hfunc, int texunit, CUtexref hTexRef);
+static CUresult (*pcuParamSetf)(CUfunction hfunc, int offset, float value);
+static CUresult (*pcuParamSeti)(CUfunction hfunc, int offset, unsigned int value);
+static CUresult (*pcuParamSetv)(CUfunction hfunc, int offset, void *ptr, unsigned int numbytes);
+static CUresult (*pcuPointerGetAttribute)(void *data, CUpointer_attribute attribute, CUdeviceptr ptr);
+static CUresult (*pcuPointerSetAttribute)(const void *value, CUpointer_attribute attribute, CUdeviceptr ptr);
+static CUresult (*pcuStreamAddCallback)(CUstream hStream, void *callback, void *userData, unsigned int flags);
+static CUresult (*pcuStreamAttachMemAsync)(CUstream hStream, CUdeviceptr dptr, size_t length, unsigned int flags);
+static CUresult (*pcuStreamCreate)(CUstream *phStream, unsigned int Flags);
+static CUresult (*pcuStreamCreateWithPriority)(CUstream *phStream, unsigned int flags, int priority);
+static CUresult (*pcuStreamDestroy)(CUstream hStream);
+static CUresult (*pcuStreamDestroy_v2)(CUstream hStream);
+static CUresult (*pcuStreamGetFlags)(CUstream hStream, unsigned int *flags);
+static CUresult (*pcuStreamGetPriority)(CUstream hStream, int *priority);
+static CUresult (*pcuStreamQuery)(CUstream hStream);
+static CUresult (*pcuStreamSynchronize)(CUstream hStream);
+static CUresult (*pcuStreamWaitEvent)(CUstream hStream, CUevent hEvent, unsigned int Flags);
+static CUresult (*pcuSurfObjectCreate)(CUsurfObject *pSurfObject, const CUDA_RESOURCE_DESC *pResDesc);
+static CUresult (*pcuSurfObjectDestroy)(CUsurfObject surfObject);
+static CUresult (*pcuSurfObjectGetResourceDesc)(CUDA_RESOURCE_DESC *pResDesc, CUsurfObject surfObject);
+static CUresult (*pcuSurfRefGetArray)(CUarray *phArray, CUsurfref hSurfRef);
+static CUresult (*pcuSurfRefSetArray)(CUsurfref hSurfRef, CUarray hArray, unsigned int Flags);
+static CUresult (*pcuTexObjectCreate)(CUtexObject *pTexObject, const CUDA_RESOURCE_DESC *pResDesc,
+                                      const CUDA_TEXTURE_DESC *pTexDesc, const CUDA_RESOURCE_VIEW_DESC *pResViewDesc);
+static CUresult (*pcuTexObjectDestroy)(CUtexObject texObject);
+static CUresult (*pcuTexObjectGetResourceDesc)(CUDA_RESOURCE_DESC *pResDesc, CUtexObject texObject);
+static CUresult (*pcuTexObjectGetResourceViewDesc)(CUDA_RESOURCE_VIEW_DESC *pResViewDesc, CUtexObject texObject);
+static CUresult (*pcuTexObjectGetTextureDesc)(CUDA_TEXTURE_DESC *pTexDesc, CUtexObject texObject);
+static CUresult (*pcuTexRefCreate)(CUtexref *pTexRef);
+static CUresult (*pcuTexRefDestroy)(CUtexref hTexRef);
+static CUresult (*pcuTexRefGetAddress)(CUdeviceptr *pdptr, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetAddressMode)(CUaddress_mode *pam, CUtexref hTexRef, int dim);
+static CUresult (*pcuTexRefGetAddress_v2)(CUdeviceptr *pdptr, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetArray)(CUarray *phArray, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetFilterMode)(CUfilter_mode *pfm, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetFlags)(unsigned int *pFlags, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetFormat)(CUarray_format *pFormat, int *pNumChannels, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetMaxAnisotropy)(int *pmaxAniso, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetMipmapFilterMode)(CUfilter_mode *pfm, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetMipmapLevelBias)(float *pbias, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetMipmapLevelClamp)(float *pminMipmapLevelClamp, float *pmaxMipmapLevelClamp, CUtexref hTexRef);
+static CUresult (*pcuTexRefGetMipmappedArray)(CUmipmappedArray *phMipmappedArray, CUtexref hTexRef);
+static CUresult (*pcuTexRefSetAddress)(size_t *ByteOffset, CUtexref hTexRef, CUdeviceptr dptr, size_t bytes);
+static CUresult (*pcuTexRefSetAddress2D)(CUtexref hTexRef, const CUDA_ARRAY_DESCRIPTOR *desc, CUdeviceptr dptr, unsigned int Pitch);
+static CUresult (*pcuTexRefSetAddress2D_v2)(CUtexref hTexRef, const CUDA_ARRAY_DESCRIPTOR *desc, CUdeviceptr dptr, unsigned int Pitch);
+static CUresult (*pcuTexRefSetAddress2D_v3)(CUtexref hTexRef, const CUDA_ARRAY_DESCRIPTOR *desc, CUdeviceptr dptr, unsigned int Pitch);
+static CUresult (*pcuTexRefSetAddressMode)(CUtexref hTexRef, int dim, CUaddress_mode am);
+static CUresult (*pcuTexRefSetAddress_v2)(size_t *ByteOffset, CUtexref hTexRef, CUdeviceptr dptr, size_t bytes);
+static CUresult (*pcuTexRefSetArray)(CUtexref hTexRef, CUarray hArray, unsigned int Flags);
+static CUresult (*pcuTexRefSetFilterMode)(CUtexref hTexRef, CUfilter_mode fm);
+static CUresult (*pcuTexRefSetFlags)(CUtexref hTexRef, unsigned int Flags);
+static CUresult (*pcuTexRefSetFormat)(CUtexref hTexRef, CUarray_format fmt, int NumPackedComponents);
+static CUresult (*pcuTexRefSetMaxAnisotropy)(CUtexref hTexRef, unsigned int maxAniso);
+static CUresult (*pcuTexRefSetMipmapFilterMode)(CUtexref hTexRef, CUfilter_mode fm);
+static CUresult (*pcuTexRefSetMipmapLevelBias)(CUtexref hTexRef, float bias);
+static CUresult (*pcuTexRefSetMipmapLevelClamp)(CUtexref hTexRef, float minMipmapLevelClamp, float maxMipmapLevelClamp);
+static CUresult (*pcuTexRefSetMipmappedArray)(CUtexref hTexRef, CUmipmappedArray hMipmappedArray, unsigned int Flags);
+
+static void *cuda_handle = NULL;
+
+static BOOL load_functions(void)
+{
+    cuda_handle = wine_dlopen("libcuda.so", RTLD_NOW, NULL, 0);
+
+    if (!cuda_handle)
+    {
+        FIXME("Wine cannot find the libcuda.so library, CUDA support disabled.\n");
+        return FALSE;
+    }
+
+    #define LOAD_FUNCPTR(f) if((p##f = wine_dlsym(cuda_handle, #f, NULL, 0)) == NULL){FIXME("Can't find symbol %s\n", #f); return FALSE;}
+
+    LOAD_FUNCPTR(cuArray3DCreate);
+    LOAD_FUNCPTR(cuArray3DCreate_v2);
+    LOAD_FUNCPTR(cuArray3DGetDescriptor);
+    LOAD_FUNCPTR(cuArray3DGetDescriptor_v2);
+    LOAD_FUNCPTR(cuArrayCreate);
+    LOAD_FUNCPTR(cuArrayCreate_v2);
+    LOAD_FUNCPTR(cuArrayDestroy);
+    LOAD_FUNCPTR(cuArrayGetDescriptor);
+    LOAD_FUNCPTR(cuArrayGetDescriptor_v2);
+    LOAD_FUNCPTR(cuCtxAttach);
+    LOAD_FUNCPTR(cuCtxCreate);
+    LOAD_FUNCPTR(cuCtxCreate_v2);
+    LOAD_FUNCPTR(cuCtxDestroy);
+    LOAD_FUNCPTR(cuCtxDestroy_v2);
+    LOAD_FUNCPTR(cuCtxDetach);
+    LOAD_FUNCPTR(cuCtxDisablePeerAccess);
+    LOAD_FUNCPTR(cuCtxEnablePeerAccess);
+    LOAD_FUNCPTR(cuCtxGetApiVersion);
+    LOAD_FUNCPTR(cuCtxGetCacheConfig);
+    LOAD_FUNCPTR(cuCtxGetCurrent);
+    LOAD_FUNCPTR(cuCtxGetDevice);
+    LOAD_FUNCPTR(cuCtxGetLimit);
+    LOAD_FUNCPTR(cuCtxGetSharedMemConfig);
+    LOAD_FUNCPTR(cuCtxGetStreamPriorityRange);
+    LOAD_FUNCPTR(cuCtxPopCurrent);
+    LOAD_FUNCPTR(cuCtxPopCurrent_v2);
+    LOAD_FUNCPTR(cuCtxPushCurrent);
+    LOAD_FUNCPTR(cuCtxPushCurrent_v2);
+    LOAD_FUNCPTR(cuCtxSetCacheConfig);
+    LOAD_FUNCPTR(cuCtxSetCurrent);
+    LOAD_FUNCPTR(cuCtxSetLimit);
+    LOAD_FUNCPTR(cuCtxSetSharedMemConfig);
+    LOAD_FUNCPTR(cuCtxSynchronize);
+    LOAD_FUNCPTR(cuDeviceCanAccessPeer);
+    LOAD_FUNCPTR(cuDeviceComputeCapability);
+    LOAD_FUNCPTR(cuDeviceGet);
+    LOAD_FUNCPTR(cuDeviceGetAttribute);
+    LOAD_FUNCPTR(cuDeviceGetByPCIBusId);
+    LOAD_FUNCPTR(cuDeviceGetCount);
+    LOAD_FUNCPTR(cuDeviceGetName);
+    LOAD_FUNCPTR(cuDeviceGetPCIBusId);
+    LOAD_FUNCPTR(cuDeviceGetProperties);
+    LOAD_FUNCPTR(cuDeviceTotalMem);
+    LOAD_FUNCPTR(cuDeviceTotalMem_v2);
+    LOAD_FUNCPTR(cuDriverGetVersion);
+    LOAD_FUNCPTR(cuEventCreate);
+    LOAD_FUNCPTR(cuEventDestroy);
+    LOAD_FUNCPTR(cuEventDestroy_v2);
+    LOAD_FUNCPTR(cuEventElapsedTime);
+    LOAD_FUNCPTR(cuEventQuery);
+    LOAD_FUNCPTR(cuEventRecord);
+    LOAD_FUNCPTR(cuEventSynchronize);
+    LOAD_FUNCPTR(cuFuncGetAttribute);
+    LOAD_FUNCPTR(cuFuncSetBlockShape);
+    LOAD_FUNCPTR(cuFuncSetCacheConfig);
+    LOAD_FUNCPTR(cuFuncSetSharedMemConfig);
+    LOAD_FUNCPTR(cuFuncSetSharedSize);
+    LOAD_FUNCPTR(cuGLCtxCreate);
+    LOAD_FUNCPTR(cuGLCtxCreate_v2);
+    LOAD_FUNCPTR(cuGLGetDevices);
+    LOAD_FUNCPTR(cuGLInit);
+    LOAD_FUNCPTR(cuGLMapBufferObject);
+    LOAD_FUNCPTR(cuGLMapBufferObjectAsync);
+    LOAD_FUNCPTR(cuGLMapBufferObjectAsync_v2);
+    LOAD_FUNCPTR(cuGLMapBufferObject_v2);
+    LOAD_FUNCPTR(cuGLRegisterBufferObject);
+    LOAD_FUNCPTR(cuGLSetBufferObjectMapFlags);
+    LOAD_FUNCPTR(cuGLUnmapBufferObject);
+    LOAD_FUNCPTR(cuGLUnmapBufferObjectAsync);
+    LOAD_FUNCPTR(cuGLUnregisterBufferObject);
+    LOAD_FUNCPTR(cuGetErrorName);
+    LOAD_FUNCPTR(cuGetErrorString);
+    LOAD_FUNCPTR(cuGetExportTable);
+    LOAD_FUNCPTR(cuGraphicsGLRegisterBuffer);
+    LOAD_FUNCPTR(cuGraphicsGLRegisterImage);
+    LOAD_FUNCPTR(cuGraphicsMapResources);
+    LOAD_FUNCPTR(cuGraphicsResourceGetMappedMipmappedArray);
+    LOAD_FUNCPTR(cuGraphicsResourceGetMappedPointer);
+    LOAD_FUNCPTR(cuGraphicsResourceGetMappedPointer_v2);
+    LOAD_FUNCPTR(cuGraphicsResourceSetMapFlags);
+    LOAD_FUNCPTR(cuGraphicsSubResourceGetMappedArray);
+    LOAD_FUNCPTR(cuGraphicsUnmapResources);
+    LOAD_FUNCPTR(cuGraphicsUnregisterResource);
+    LOAD_FUNCPTR(cuInit);
+    LOAD_FUNCPTR(cuIpcCloseMemHandle);
+    LOAD_FUNCPTR(cuIpcGetEventHandle);
+    LOAD_FUNCPTR(cuIpcGetMemHandle);
+    LOAD_FUNCPTR(cuIpcOpenEventHandle);
+    LOAD_FUNCPTR(cuIpcOpenMemHandle);
+    LOAD_FUNCPTR(cuLaunch);
+    LOAD_FUNCPTR(cuLaunchGrid);
+    LOAD_FUNCPTR(cuLaunchGridAsync);
+    LOAD_FUNCPTR(cuLaunchKernel);
+    LOAD_FUNCPTR(cuLinkAddData);
+    LOAD_FUNCPTR(cuLinkComplete);
+    LOAD_FUNCPTR(cuLinkCreate);
+    LOAD_FUNCPTR(cuLinkDestroy);
+    LOAD_FUNCPTR(cuMemAlloc);
+    LOAD_FUNCPTR(cuMemAllocHost);
+    LOAD_FUNCPTR(cuMemAllocHost_v2);
+    LOAD_FUNCPTR(cuMemAllocManaged);
+    LOAD_FUNCPTR(cuMemAllocPitch);
+    LOAD_FUNCPTR(cuMemAllocPitch_v2);
+    LOAD_FUNCPTR(cuMemAlloc_v2);
+    LOAD_FUNCPTR(cuMemFree);
+    LOAD_FUNCPTR(cuMemFreeHost);
+    LOAD_FUNCPTR(cuMemFree_v2);
+    LOAD_FUNCPTR(cuMemGetAddressRange);
+    LOAD_FUNCPTR(cuMemGetAddressRange_v2);
+    LOAD_FUNCPTR(cuMemGetInfo);
+    LOAD_FUNCPTR(cuMemGetInfo_v2);
+    LOAD_FUNCPTR(cuMemHostAlloc);
+    LOAD_FUNCPTR(cuMemHostGetDevicePointer);
+    LOAD_FUNCPTR(cuMemHostGetDevicePointer_v2);
+    LOAD_FUNCPTR(cuMemHostGetFlags);
+    LOAD_FUNCPTR(cuMemHostRegister);
+    LOAD_FUNCPTR(cuMemHostUnregister);
+    LOAD_FUNCPTR(cuMemcpy);
+    LOAD_FUNCPTR(cuMemcpy2D);
+    LOAD_FUNCPTR(cuMemcpy2DAsync);
+    LOAD_FUNCPTR(cuMemcpy2DAsync_v2);
+    LOAD_FUNCPTR(cuMemcpy2DUnaligned);
+    LOAD_FUNCPTR(cuMemcpy2DUnaligned_v2);
+    LOAD_FUNCPTR(cuMemcpy2D_v2);
+    LOAD_FUNCPTR(cuMemcpy3D);
+    LOAD_FUNCPTR(cuMemcpy3DAsync);
+    LOAD_FUNCPTR(cuMemcpy3DAsync_v2);
+    LOAD_FUNCPTR(cuMemcpy3DPeer);
+    LOAD_FUNCPTR(cuMemcpy3DPeerAsync);
+    LOAD_FUNCPTR(cuMemcpy3D_v2);
+    LOAD_FUNCPTR(cuMemcpyAsync);
+    LOAD_FUNCPTR(cuMemcpyAtoA);
+    LOAD_FUNCPTR(cuMemcpyAtoA_v2);
+    LOAD_FUNCPTR(cuMemcpyAtoD);
+    LOAD_FUNCPTR(cuMemcpyAtoD_v2);
+    LOAD_FUNCPTR(cuMemcpyAtoH);
+    LOAD_FUNCPTR(cuMemcpyAtoHAsync);
+    LOAD_FUNCPTR(cuMemcpyAtoHAsync_v2);
+    LOAD_FUNCPTR(cuMemcpyAtoH_v2);
+    LOAD_FUNCPTR(cuMemcpyDtoA);
+    LOAD_FUNCPTR(cuMemcpyDtoA_v2);
+    LOAD_FUNCPTR(cuMemcpyDtoD);
+    LOAD_FUNCPTR(cuMemcpyDtoDAsync);
+    LOAD_FUNCPTR(cuMemcpyDtoDAsync_v2);
+    LOAD_FUNCPTR(cuMemcpyDtoD_v2);
+    LOAD_FUNCPTR(cuMemcpyDtoH);
+    LOAD_FUNCPTR(cuMemcpyDtoHAsync);
+    LOAD_FUNCPTR(cuMemcpyDtoHAsync_v2);
+    LOAD_FUNCPTR(cuMemcpyDtoH_v2);
+    LOAD_FUNCPTR(cuMemcpyHtoA);
+    LOAD_FUNCPTR(cuMemcpyHtoAAsync);
+    LOAD_FUNCPTR(cuMemcpyHtoAAsync_v2);
+    LOAD_FUNCPTR(cuMemcpyHtoA_v2);
+    LOAD_FUNCPTR(cuMemcpyHtoD);
+    LOAD_FUNCPTR(cuMemcpyHtoDAsync);
+    LOAD_FUNCPTR(cuMemcpyHtoDAsync_v2);
+    LOAD_FUNCPTR(cuMemcpyHtoD_v2);
+    LOAD_FUNCPTR(cuMemcpyPeer);
+    LOAD_FUNCPTR(cuMemcpyPeerAsync);
+    LOAD_FUNCPTR(cuMemsetD16);
+    LOAD_FUNCPTR(cuMemsetD16Async);
+    LOAD_FUNCPTR(cuMemsetD16_v2);
+    LOAD_FUNCPTR(cuMemsetD2D16);
+    LOAD_FUNCPTR(cuMemsetD2D16Async);
+    LOAD_FUNCPTR(cuMemsetD2D16_v2);
+    LOAD_FUNCPTR(cuMemsetD2D32);
+    LOAD_FUNCPTR(cuMemsetD2D32Async);
+    LOAD_FUNCPTR(cuMemsetD2D32_v2);
+    LOAD_FUNCPTR(cuMemsetD2D8);
+    LOAD_FUNCPTR(cuMemsetD2D8Async);
+    LOAD_FUNCPTR(cuMemsetD2D8_v2);
+    LOAD_FUNCPTR(cuMemsetD32);
+    LOAD_FUNCPTR(cuMemsetD32Async);
+    LOAD_FUNCPTR(cuMemsetD32_v2);
+    LOAD_FUNCPTR(cuMemsetD8);
+    LOAD_FUNCPTR(cuMemsetD8Async);
+    LOAD_FUNCPTR(cuMemsetD8_v2);
+    LOAD_FUNCPTR(cuMipmappedArrayCreate);
+    LOAD_FUNCPTR(cuMipmappedArrayDestroy);
+    LOAD_FUNCPTR(cuMipmappedArrayGetLevel);
+    LOAD_FUNCPTR(cuModuleGetFunction);
+    LOAD_FUNCPTR(cuModuleGetGlobal);
+    LOAD_FUNCPTR(cuModuleGetGlobal_v2);
+    LOAD_FUNCPTR(cuModuleGetSurfRef);
+    LOAD_FUNCPTR(cuModuleGetTexRef);
+    LOAD_FUNCPTR(cuModuleLoadData);
+    LOAD_FUNCPTR(cuModuleLoadDataEx);
+    LOAD_FUNCPTR(cuModuleLoadFatBinary);
+    LOAD_FUNCPTR(cuModuleUnload);
+    LOAD_FUNCPTR(cuParamSetSize);
+    LOAD_FUNCPTR(cuParamSetTexRef);
+    LOAD_FUNCPTR(cuParamSetf);
+    LOAD_FUNCPTR(cuParamSeti);
+    LOAD_FUNCPTR(cuParamSetv);
+    LOAD_FUNCPTR(cuPointerGetAttribute);
+    LOAD_FUNCPTR(cuPointerSetAttribute);
+    LOAD_FUNCPTR(cuStreamAddCallback);
+    LOAD_FUNCPTR(cuStreamAttachMemAsync);
+    LOAD_FUNCPTR(cuStreamCreate);
+    LOAD_FUNCPTR(cuStreamCreateWithPriority);
+    LOAD_FUNCPTR(cuStreamDestroy);
+    LOAD_FUNCPTR(cuStreamDestroy_v2);
+    LOAD_FUNCPTR(cuStreamGetFlags);
+    LOAD_FUNCPTR(cuStreamGetPriority);
+    LOAD_FUNCPTR(cuStreamQuery);
+    LOAD_FUNCPTR(cuStreamSynchronize);
+    LOAD_FUNCPTR(cuStreamWaitEvent);
+    LOAD_FUNCPTR(cuSurfObjectCreate);
+    LOAD_FUNCPTR(cuSurfObjectDestroy);
+    LOAD_FUNCPTR(cuSurfObjectGetResourceDesc);
+    LOAD_FUNCPTR(cuSurfRefGetArray);
+    LOAD_FUNCPTR(cuSurfRefSetArray);
+    LOAD_FUNCPTR(cuTexObjectCreate);
+    LOAD_FUNCPTR(cuTexObjectDestroy);
+    LOAD_FUNCPTR(cuTexObjectGetResourceDesc);
+    LOAD_FUNCPTR(cuTexObjectGetResourceViewDesc);
+    LOAD_FUNCPTR(cuTexObjectGetTextureDesc);
+    LOAD_FUNCPTR(cuTexRefCreate);
+    LOAD_FUNCPTR(cuTexRefDestroy);
+    LOAD_FUNCPTR(cuTexRefGetAddress);
+    LOAD_FUNCPTR(cuTexRefGetAddressMode);
+    LOAD_FUNCPTR(cuTexRefGetAddress_v2);
+    LOAD_FUNCPTR(cuTexRefGetArray);
+    LOAD_FUNCPTR(cuTexRefGetFilterMode);
+    LOAD_FUNCPTR(cuTexRefGetFlags);
+    LOAD_FUNCPTR(cuTexRefGetFormat);
+    LOAD_FUNCPTR(cuTexRefGetMaxAnisotropy);
+    LOAD_FUNCPTR(cuTexRefGetMipmapFilterMode);
+    LOAD_FUNCPTR(cuTexRefGetMipmapLevelBias);
+    LOAD_FUNCPTR(cuTexRefGetMipmapLevelClamp);
+    LOAD_FUNCPTR(cuTexRefGetMipmappedArray);
+    LOAD_FUNCPTR(cuTexRefSetAddress);
+    LOAD_FUNCPTR(cuTexRefSetAddress2D);
+    LOAD_FUNCPTR(cuTexRefSetAddress2D_v2);
+    LOAD_FUNCPTR(cuTexRefSetAddress2D_v3);
+    LOAD_FUNCPTR(cuTexRefSetAddressMode);
+    LOAD_FUNCPTR(cuTexRefSetAddress_v2);
+    LOAD_FUNCPTR(cuTexRefSetArray)
+    LOAD_FUNCPTR(cuTexRefSetFilterMode);
+    LOAD_FUNCPTR(cuTexRefSetFlags);
+    LOAD_FUNCPTR(cuTexRefSetFormat);
+    LOAD_FUNCPTR(cuTexRefSetMaxAnisotropy);
+    LOAD_FUNCPTR(cuTexRefSetMipmapFilterMode);
+    LOAD_FUNCPTR(cuTexRefSetMipmapLevelBias);
+    LOAD_FUNCPTR(cuTexRefSetMipmapLevelClamp);
+    LOAD_FUNCPTR(cuTexRefSetMipmappedArray);
+
+    #undef LOAD_FUNCPTR
+
+    return TRUE;
+}
+
+CUresult WINAPI wine_cuArray3DCreate(CUarray *pHandle, const CUDA_ARRAY3D_DESCRIPTOR *pAllocateArray)
+{
+    TRACE("(%p, %p)\n", pHandle, pAllocateArray);
+    return pcuArray3DCreate(pHandle, pAllocateArray);
+}
+
+CUresult WINAPI wine_cuArray3DCreate_v2(CUarray *pHandle, const CUDA_ARRAY3D_DESCRIPTOR *pAllocateArray)
+{
+    TRACE("(%p, %p)\n", pHandle, pAllocateArray);
+    return pcuArray3DCreate_v2(pHandle, pAllocateArray);
+}
+
+CUresult WINAPI wine_cuArray3DGetDescriptor(CUDA_ARRAY3D_DESCRIPTOR *pArrayDescriptor, CUarray hArray)
+{
+    TRACE("(%p, %p)\n", pArrayDescriptor, hArray);
+    return pcuArray3DGetDescriptor(pArrayDescriptor, hArray);
+}
+
+CUresult WINAPI wine_cuArray3DGetDescriptor_v2(CUDA_ARRAY3D_DESCRIPTOR *pArrayDescriptor, CUarray hArray)
+{
+    TRACE("(%p, %p)\n", pArrayDescriptor, hArray);
+    return pcuArray3DGetDescriptor_v2(pArrayDescriptor, hArray);
+}
+
+CUresult WINAPI wine_cuArrayCreate(CUarray *pHandle, const CUDA_ARRAY_DESCRIPTOR *pAllocateArray)
+{
+    TRACE("(%p, %p)\n", pHandle, pAllocateArray);
+    return pcuArrayCreate(pHandle, pAllocateArray);
+}
+
+CUresult WINAPI wine_cuArrayCreate_v2(CUarray *pHandle, const CUDA_ARRAY_DESCRIPTOR *pAllocateArray)
+{
+    TRACE("(%p, %p)\n", pHandle, pAllocateArray);
+    return pcuArrayCreate_v2(pHandle, pAllocateArray);
+}
+
+CUresult WINAPI wine_cuArrayDestroy(CUarray hArray)
+{
+    TRACE("(%p)\n", hArray);
+    return pcuArrayDestroy(hArray);
+}
+
+CUresult WINAPI wine_cuArrayGetDescriptor(CUDA_ARRAY_DESCRIPTOR *pArrayDescriptor, CUarray hArray)
+{
+    TRACE("(%p, %p)\n", pArrayDescriptor, hArray);
+    return pcuArrayGetDescriptor(pArrayDescriptor, hArray);
+}
+
+CUresult WINAPI wine_cuArrayGetDescriptor_v2(CUDA_ARRAY_DESCRIPTOR *pArrayDescriptor, CUarray hArray)
+{
+    TRACE("(%p, %p)\n", pArrayDescriptor, hArray);
+    return pcuArrayGetDescriptor_v2(pArrayDescriptor, hArray);
+}
+
+CUresult WINAPI wine_cuCtxAttach(CUcontext *pctx, unsigned int flags)
+{
+    TRACE("(%p, %u)\n", pctx, flags);
+    return pcuCtxAttach(pctx, flags);
+}
+
+CUresult WINAPI wine_cuCtxCreate(CUcontext *pctx, unsigned int flags, CUdevice dev)
+{
+    TRACE("(%p, %u, %u)\n", pctx, flags, dev);
+    return pcuCtxCreate(pctx, flags, dev);
+}
+
+CUresult WINAPI wine_cuCtxCreate_v2(CUcontext *pctx, unsigned int flags, CUdevice dev)
+{
+    TRACE("(%p, %u, %u)\n", pctx, flags, dev);
+    return pcuCtxCreate_v2(pctx, flags, dev);
+}
+
+CUresult WINAPI wine_cuCtxDestroy(CUcontext ctx)
+{
+    TRACE("(%p)\n", ctx);
+    return pcuCtxDestroy(ctx);
+}
+
+CUresult WINAPI wine_cuCtxDestroy_v2(CUcontext ctx)
+{
+    TRACE("(%p)\n", ctx);
+    return pcuCtxDestroy_v2(ctx);
+}
+
+CUresult WINAPI wine_cuCtxDetach(CUcontext ctx)
+{
+    TRACE("(%p)\n", ctx);
+    return pcuCtxDetach(ctx);
+}
+
+CUresult WINAPI wine_cuCtxDisablePeerAccess(CUcontext peerContext)
+{
+    TRACE("(%p)\n", peerContext);
+    return pcuCtxDisablePeerAccess(peerContext);
+}
+
+CUresult WINAPI wine_cuCtxEnablePeerAccess(CUcontext peerContext, unsigned int Flags)
+{
+    TRACE("(%p, %u)\n", peerContext, Flags);
+    return pcuCtxEnablePeerAccess(peerContext, Flags);
+}
+
+CUresult WINAPI wine_cuCtxGetApiVersion(CUcontext ctx, unsigned int *version)
+{
+    TRACE("(%p, %p)\n", ctx, version);
+    return pcuCtxGetApiVersion(ctx, version);
+}
+
+CUresult WINAPI wine_cuCtxGetCacheConfig(CUfunc_cache *pconfig)
+{
+    TRACE("(%p)\n", pconfig);
+    return pcuCtxGetCacheConfig(pconfig);
+}
+
+CUresult WINAPI wine_cuCtxGetCurrent(CUcontext *pctx)
+{
+    TRACE("(%p)\n", pctx);
+    return pcuCtxGetCurrent(pctx);
+}
+
+CUresult WINAPI wine_cuCtxGetDevice(CUdevice *device)
+{
+    TRACE("(%p)\n", device);
+    return pcuCtxGetDevice(device);
+}
+
+CUresult WINAPI wine_cuCtxGetLimit(size_t *pvalue, CUlimit limit)
+{
+    TRACE("(%p, %d)\n", pvalue, limit);
+    return pcuCtxGetLimit(pvalue, limit);
+}
+
+CUresult WINAPI wine_cuCtxGetSharedMemConfig(CUsharedconfig *pConfig)
+{
+    TRACE("(%p)\n", pConfig);
+    return pcuCtxGetSharedMemConfig(pConfig);
+}
+
+CUresult WINAPI wine_cuCtxGetStreamPriorityRange(int *leastPriority, int *greatestPriority)
+{
+    TRACE("(%p, %p)\n", leastPriority, greatestPriority);
+    return pcuCtxGetStreamPriorityRange(leastPriority, greatestPriority);
+}
+
+CUresult WINAPI wine_cuCtxPopCurrent(CUcontext *pctx)
+{
+    TRACE("(%p)\n", pctx);
+    return pcuCtxPopCurrent(pctx);
+}
+
+CUresult WINAPI wine_cuCtxPopCurrent_v2(CUcontext *pctx)
+{
+    TRACE("(%p)\n", pctx);
+    return pcuCtxPopCurrent_v2(pctx);
+}
+
+CUresult WINAPI wine_cuCtxPushCurrent(CUcontext ctx)
+{
+    TRACE("(%p)\n", ctx);
+    return pcuCtxPushCurrent(ctx);
+}
+
+CUresult WINAPI wine_cuCtxPushCurrent_v2(CUcontext ctx)
+{
+    TRACE("(%p)\n", ctx);
+    return pcuCtxPushCurrent_v2(ctx);
+}
+
+CUresult WINAPI wine_cuCtxSetCacheConfig(CUfunc_cache config)
+{
+    TRACE("(%d)\n", config);
+    return pcuCtxSetCacheConfig(config);
+}
+
+CUresult WINAPI wine_cuCtxSetCurrent(CUcontext ctx)
+{
+    TRACE("(%p)\n", ctx);
+    return pcuCtxSetCurrent(ctx);
+}
+
+CUresult WINAPI wine_cuCtxSetLimit(CUlimit limit, size_t value)
+{
+    TRACE("(%d, %lu)\n", limit, (SIZE_T)value);
+    return pcuCtxSetLimit(limit, value);
+}
+
+CUresult WINAPI wine_cuCtxSetSharedMemConfig(CUsharedconfig config)
+{
+    TRACE("(%d)\n", config);
+    return pcuCtxSetSharedMemConfig(config);
+}
+
+CUresult WINAPI wine_cuCtxSynchronize(void)
+{
+    TRACE("()\n");
+    return pcuCtxSynchronize();
+}
+
+CUresult WINAPI wine_cuDeviceCanAccessPeer(int *canAccessPeer, CUdevice dev, CUdevice peerDev)
+{
+    TRACE("(%p, %u, %u)\n", canAccessPeer, dev, peerDev);
+    return pcuDeviceCanAccessPeer(canAccessPeer, dev, peerDev);
+}
+
+CUresult WINAPI wine_cuDeviceComputeCapability(int *major, int *minor, CUdevice dev)
+{
+    TRACE("(%p, %p, %d)\n", major, minor, dev);
+    return pcuDeviceComputeCapability(major, minor, dev);
+}
+
+CUresult WINAPI wine_cuDeviceGet(CUdevice *device, int ordinal)
+{
+    TRACE("(%p, %d)\n", device, ordinal);
+    return pcuDeviceGet(device, ordinal);
+}
+
+CUresult WINAPI wine_cuDeviceGetAttribute(int *pi, CUdevice_attribute attrib, CUdevice dev)
+{
+    TRACE("(%p, %d, %d)\n", pi, attrib, dev);
+    return pcuDeviceGetAttribute(pi, attrib, dev);;
+}
+
+CUresult WINAPI wine_cuDeviceGetByPCIBusId(CUdevice *dev, const char *pciBusId)
+{
+    TRACE("(%p, %s)\n", dev, pciBusId);
+    return pcuDeviceGetByPCIBusId(dev, pciBusId);
+}
+
+CUresult WINAPI wine_cuDeviceGetCount(int *count)
+{
+    TRACE("(%p)\n", count);
+    return pcuDeviceGetCount(count);
+}
+
+CUresult WINAPI wine_cuDeviceGetName(char *name, int len, CUdevice dev)
+{
+    TRACE("(%p, %d, %d)\n", name, len, dev);
+    return pcuDeviceGetName(name, len, dev);
+}
+
+CUresult WINAPI wine_cuDeviceGetPCIBusId(char *pciBusId, int len, CUdevice dev)
+{
+    TRACE("(%p, %d, %d)\n", pciBusId, len, dev);
+    return pcuDeviceGetPCIBusId(pciBusId, len, dev);
+}
+
+CUresult WINAPI wine_cuDeviceGetProperties(CUdevprop *prop, CUdevice dev)
+{
+    TRACE("(%p, %d)\n", prop, dev);
+    return pcuDeviceGetProperties(prop, dev);
+}
+
+CUresult WINAPI wine_cuDeviceTotalMem(size_t *bytes, CUdevice dev)
+{
+    TRACE("(%p, %d)\n", bytes, dev);
+    return pcuDeviceTotalMem(bytes, dev);
+}
+
+CUresult WINAPI wine_cuDeviceTotalMem_v2(size_t *bytes, CUdevice dev)
+{
+    TRACE("(%p, %d)\n", bytes, dev);
+    return pcuDeviceTotalMem_v2(bytes, dev);
+}
+
+CUresult WINAPI wine_cuDriverGetVersion(int *version)
+{
+    TRACE("(%p)\n", version);
+    return pcuDriverGetVersion(version);
+}
+
+CUresult WINAPI wine_cuEventCreate(CUevent *phEvent, unsigned int Flags)
+{
+    TRACE("(%p, %u)\n", phEvent, Flags);
+    return pcuEventCreate(phEvent, Flags);
+}
+
+CUresult WINAPI wine_cuEventDestroy(CUevent hEvent)
+{
+    TRACE("(%p)\n", hEvent);
+    return pcuEventDestroy(hEvent);
+}
+
+CUresult WINAPI wine_cuEventDestroy_v2(CUevent hEvent)
+{
+    TRACE("(%p)\n", hEvent);
+    return pcuEventDestroy_v2(hEvent);
+}
+
+CUresult WINAPI wine_cuEventElapsedTime(float *pMilliseconds, CUevent hStart, CUevent hEnd)
+{
+    TRACE("(%p, %p, %p)\n", pMilliseconds, hStart, hEnd);
+    return pcuEventElapsedTime(pMilliseconds, hStart, hEnd);
+}
+
+CUresult WINAPI wine_cuEventQuery(CUevent hEvent)
+{
+    TRACE("(%p)\n", hEvent);
+    return pcuEventQuery(hEvent);
+}
+
+CUresult WINAPI wine_cuEventRecord(CUevent hEvent, CUstream hStream)
+{
+    TRACE("(%p, %p)\n", hEvent, hStream);
+    return pcuEventRecord(hEvent, hStream);
+}
+
+CUresult WINAPI wine_cuEventSynchronize(CUevent hEvent)
+{
+    TRACE("(%p)\n", hEvent);
+    return pcuEventSynchronize(hEvent);
+}
+
+CUresult WINAPI wine_cuFuncGetAttribute(int *pi, CUfunction_attribute attrib, CUfunction hfunc)
+{
+    TRACE("(%p, %d, %p)\n", pi, attrib, hfunc);
+    return pcuFuncGetAttribute(pi, attrib, hfunc);
+}
+
+CUresult WINAPI wine_cuFuncSetBlockShape(CUfunction hfunc, int x, int y, int z)
+{
+    TRACE("(%p, %d, %d, %d)\n", hfunc, x, y, z);
+    return pcuFuncSetBlockShape(hfunc, x, y, z);
+}
+
+CUresult WINAPI wine_cuFuncSetCacheConfig(CUfunction hfunc, CUfunc_cache config)
+{
+    TRACE("(%p, %d)\n", hfunc, config);
+    return pcuFuncSetCacheConfig(hfunc, config);
+}
+
+CUresult WINAPI wine_cuFuncSetSharedMemConfig(CUfunction hfunc, CUsharedconfig config)
+{
+    TRACE("(%p, %d)\n", hfunc, config);
+    return pcuFuncSetSharedMemConfig(hfunc, config);
+}
+
+CUresult WINAPI wine_cuFuncSetSharedSize(CUfunction hfunc, unsigned int bytes)
+{
+    TRACE("(%p, %u)\n", hfunc, bytes);
+    return pcuFuncSetSharedSize(hfunc, bytes);
+}
+
+CUresult WINAPI wine_cuGLCtxCreate(CUcontext *pCtx, unsigned int Flags, CUdevice device)
+{
+    TRACE("(%p, %u, %d)\n", pCtx, Flags, device);
+    return pcuGLCtxCreate(pCtx, Flags, device);
+}
+
+CUresult WINAPI wine_cuGLCtxCreate_v2(CUcontext *pCtx, unsigned int Flags, CUdevice device)
+{
+    TRACE("(%p, %u, %d)\n", pCtx, Flags, device);
+    return pcuGLCtxCreate_v2(pCtx, Flags, device);
+}
+
+CUresult WINAPI wine_cuGLGetDevices(unsigned int *pCudaDeviceCount, CUdevice *pCudaDevices,
+                                    unsigned int cudaDeviceCount, CUGLDeviceList deviceList)
+{
+    TRACE("(%p, %p, %u, %d)\n", pCudaDeviceCount, pCudaDevices, cudaDeviceCount, deviceList);
+    return pcuGLGetDevices(pCudaDeviceCount, pCudaDevices, cudaDeviceCount, deviceList);
+}
+
+CUresult WINAPI wine_cuGLInit(void)
+{
+    TRACE("()\n");
+    return pcuGLInit();
+}
+
+CUresult WINAPI wine_cuGLMapBufferObject(CUdeviceptr *dptr, size_t *size, GLuint buffer)
+{
+    TRACE("(%p, %p, %u)\n", dptr, size, buffer);
+    return pcuGLMapBufferObject(dptr, size, buffer);
+}
+
+CUresult WINAPI wine_cuGLMapBufferObjectAsync(CUdeviceptr *dptr, size_t *size, GLuint buffer, CUstream hStream)
+{
+    TRACE("(%p, %p, %u, %p)\n", dptr, size,  buffer, hStream);
+    return pcuGLMapBufferObjectAsync(dptr, size,  buffer, hStream);
+}
+
+CUresult WINAPI wine_cuGLMapBufferObjectAsync_v2(CUdeviceptr *dptr, size_t *size, GLuint buffer, CUstream hStream)
+{
+    TRACE("(%p, %p, %u, %p)\n", dptr, size,  buffer, hStream);
+    return pcuGLMapBufferObjectAsync_v2(dptr, size,  buffer, hStream);
+}
+
+CUresult WINAPI wine_cuGLMapBufferObject_v2(CUdeviceptr *dptr, size_t *size, GLuint buffer)
+{
+    TRACE("(%p, %p, %u)\n", dptr, size, buffer);
+    return pcuGLMapBufferObject_v2(dptr, size, buffer);
+}
+
+CUresult WINAPI wine_cuGLRegisterBufferObject(GLuint buffer)
+{
+    TRACE("(%u)\n", buffer);
+    return pcuGLRegisterBufferObject(buffer);
+}
+
+CUresult WINAPI wine_cuGLSetBufferObjectMapFlags(GLuint buffer, unsigned int Flags)
+{
+    TRACE("(%u, %u)\n", buffer, Flags);
+    return pcuGLSetBufferObjectMapFlags(buffer, Flags);
+}
+
+CUresult WINAPI wine_cuGLUnmapBufferObject(GLuint buffer)
+{
+    TRACE("(%u)\n", buffer);
+    return pcuGLUnmapBufferObject(buffer);
+}
+
+CUresult WINAPI wine_cuGLUnmapBufferObjectAsync(GLuint buffer, CUstream hStream)
+{
+    TRACE("(%u, %p)\n", buffer, hStream);
+    return pcuGLUnmapBufferObjectAsync(buffer, hStream);
+}
+
+CUresult WINAPI wine_cuGLUnregisterBufferObject(GLuint buffer)
+{
+    TRACE("(%u)\n", buffer);
+    return pcuGLUnregisterBufferObject(buffer);
+}
+
+CUresult WINAPI wine_cuGetErrorName(CUresult error, const char **pStr)
+{
+    TRACE("(%d, %p)\n", error, pStr);
+    return pcuGetErrorName(error, pStr);
+}
+
+CUresult WINAPI wine_cuGetErrorString(CUresult error, const char **pStr)
+{
+    TRACE("(%d, %p)\n", error, pStr);
+    return pcuGetErrorString(error, pStr);
+}
+
+CUresult WINAPI wine_cuGetExportTable(const void **table, const CUuuid *id)
+{
+    const void* orig_table = NULL;
+    CUresult ret;
+
+    TRACE("(%p, %p)\n", table, id);
+
+    ret = pcuGetExportTable(&orig_table, id);
+    return cuda_get_table(table, id, orig_table, ret);
+}
+
+CUresult WINAPI wine_cuGraphicsGLRegisterBuffer(CUgraphicsResource *pCudaResource, GLuint buffer, unsigned int Flags)
+{
+    TRACE("(%p, %u, %u)\n", pCudaResource, buffer, Flags);
+    return pcuGraphicsGLRegisterBuffer(pCudaResource, buffer, Flags);
+}
+
+CUresult WINAPI wine_cuGraphicsGLRegisterImage(CUgraphicsResource *pCudaResource, GLuint image, GLenum target, unsigned int Flags)
+{
+    TRACE("(%p, %d, %d, %d)\n", pCudaResource, image, target, Flags);
+    return pcuGraphicsGLRegisterImage(pCudaResource, image, target, Flags);
+}
+
+CUresult WINAPI wine_cuGraphicsMapResources(unsigned int count, CUgraphicsResource *resources, CUstream hStream)
+{
+    TRACE("(%u, %p, %p)\n", count, resources, hStream);
+    return pcuGraphicsMapResources(count, resources, hStream);
+}
+
+CUresult WINAPI wine_cuGraphicsResourceGetMappedMipmappedArray(CUmipmappedArray *pMipmappedArray, CUgraphicsResource resource)
+{
+    TRACE("(%p, %p)\n", pMipmappedArray, resource);
+    return pcuGraphicsResourceGetMappedMipmappedArray(pMipmappedArray, resource);
+}
+
+CUresult WINAPI wine_cuGraphicsResourceGetMappedPointer(CUdeviceptr *pDevPtr, size_t *pSize, CUgraphicsResource resource)
+{
+    TRACE("(%p, %p, %p)\n", pDevPtr, pSize, resource);
+    return pcuGraphicsResourceGetMappedPointer(pDevPtr, pSize, resource);
+}
+
+CUresult WINAPI wine_cuGraphicsResourceGetMappedPointer_v2(CUdeviceptr *pDevPtr, size_t *pSize, CUgraphicsResource resource)
+{
+    TRACE("(%p, %p, %p)\n", pDevPtr, pSize, resource);
+    return pcuGraphicsResourceGetMappedPointer_v2(pDevPtr, pSize, resource);
+}
+
+CUresult WINAPI wine_cuGraphicsResourceSetMapFlags(CUgraphicsResource resource, unsigned int flags)
+{
+    TRACE("(%p, %u)\n", resource, flags);
+    return pcuGraphicsResourceSetMapFlags(resource, flags);
+}
+
+CUresult WINAPI wine_cuGraphicsSubResourceGetMappedArray(CUarray *pArray, CUgraphicsResource resource,
+                                                         unsigned int arrayIndex, unsigned int mipLevel)
+{
+    TRACE("(%p, %p, %u, %u)\n", pArray, resource, arrayIndex, mipLevel);
+    return pcuGraphicsSubResourceGetMappedArray(pArray, resource, arrayIndex, mipLevel);
+}
+
+CUresult WINAPI wine_cuGraphicsUnmapResources(unsigned int count, CUgraphicsResource *resources, CUstream hStream)
+{
+    TRACE("(%u, %p, %p)\n", count, resources, hStream);
+    return pcuGraphicsUnmapResources(count, resources, hStream);
+}
+
+CUresult WINAPI wine_cuGraphicsUnregisterResource(CUgraphicsResource resource)
+{
+    TRACE("(%p)\n", resource);
+    return pcuGraphicsUnregisterResource(resource);
+}
+
+CUresult WINAPI wine_cuInit(unsigned int flags)
+{
+    TRACE("(%d)\n", flags);
+    return pcuInit(flags);
+}
+
+CUresult WINAPI wine_cuIpcCloseMemHandle(CUdeviceptr dptr)
+{
+    TRACE("(" DEV_PTR ")\n", dptr);
+    return pcuIpcCloseMemHandle(dptr);
+}
+
+CUresult WINAPI wine_cuIpcGetEventHandle(CUipcEventHandle *pHandle, CUevent event)
+{
+    TRACE("(%p, %p)\n", pHandle, event);
+    return pcuIpcGetEventHandle(pHandle, event);
+}
+
+CUresult WINAPI wine_cuIpcGetMemHandle(CUipcMemHandle *pHandle, CUdeviceptr dptr)
+{
+    TRACE("(%p, " DEV_PTR ")\n", pHandle, dptr);
+    return pcuIpcGetMemHandle(pHandle, dptr);
+}
+
+CUresult WINAPI wine_cuIpcOpenEventHandle(CUevent *phEvent, CUipcEventHandle handle)
+{
+    TRACE("(%p, %p)\n", phEvent, &handle); /* FIXME */
+    return pcuIpcOpenEventHandle(phEvent, handle);
+}
+
+CUresult WINAPI wine_cuIpcOpenMemHandle(CUdeviceptr *pdptr, CUipcMemHandle handle, unsigned int Flags)
+{
+    TRACE("(%p, %p, %u)\n", pdptr, &handle, Flags); /* FIXME */
+    return pcuIpcOpenMemHandle(pdptr, handle, Flags);
+}
+
+CUresult WINAPI wine_cuLaunch(CUfunction f)
+{
+    TRACE("(%p)\n", f);
+    return pcuLaunch(f);
+}
+
+CUresult WINAPI wine_cuLaunchGrid(CUfunction f, int grid_width, int grid_height)
+{
+    TRACE("(%p, %d, %d)\n", f, grid_width, grid_height);
+    return pcuLaunchGrid(f, grid_width, grid_height);
+}
+
+CUresult WINAPI wine_cuLaunchGridAsync(CUfunction f, int grid_width, int grid_height, CUstream hStream)
+{
+    TRACE("(%p, %d, %d, %p)\n", f, grid_width, grid_height, hStream);
+    return pcuLaunchGridAsync(f, grid_width, grid_height, hStream);
+}
+
+CUresult WINAPI wine_cuLaunchKernel(CUfunction f, unsigned int gridDimX, unsigned int gridDimY, unsigned int gridDimZ,
+                                    unsigned int blockDimX, unsigned int blockDimY, unsigned int blockDimZ,
+                                    unsigned int sharedMemBytes, CUstream hStream, void **kernelParams, void **extra)
+{
+    TRACE("(%p, %u, %u, %u, %u, %u, %u, %u, %p, %p, %p),\n", f, gridDimX, gridDimY, gridDimZ, blockDimX,
+          blockDimY, blockDimZ, sharedMemBytes, hStream, kernelParams, extra);
+
+    return pcuLaunchKernel(f, gridDimX, gridDimY, gridDimZ, blockDimX, blockDimY, blockDimZ, sharedMemBytes,
+                           hStream, kernelParams, extra);
+}
+
+CUresult WINAPI wine_cuLinkAddData(CUlinkState state, CUjitInputType type, void *data, size_t size, const char *name,
+                                   unsigned int numOptions, CUjit_option *options, void **optionValues)
+{
+    TRACE("(%p, %d, %p, %lu, %s, %u, %p, %p)\n", state, type, data, (SIZE_T)size, name, numOptions, options, optionValues);
+    return pcuLinkAddData(state, type, data, size, name, numOptions, options, optionValues);
+}
+
+CUresult WINAPI wine_cuLinkComplete(CUlinkState state, void **cubinOut, size_t *sizeOut)
+{
+    TRACE("(%p, %p, %p)\n", state, cubinOut, sizeOut);
+    return pcuLinkComplete(state, cubinOut, sizeOut);
+}
+
+CUresult WINAPI wine_cuLinkCreate(unsigned int numOptions, CUjit_option *options,
+                                  void **optionValues, CUlinkState *stateOut)
+{
+    TRACE("(%u, %p, %p, %p)\n", numOptions, options, optionValues, stateOut);
+    return pcuLinkCreate(numOptions, options, optionValues, stateOut);
+}
+
+CUresult WINAPI wine_cuLinkDestroy(CUlinkState state)
+{
+    TRACE("(%p)\n", state);
+    return pcuLinkDestroy(state);
+}
+
+CUresult WINAPI wine_cuMemAlloc(CUdeviceptr *dptr, unsigned int bytesize)
+{
+    TRACE("(%p, %u)\n", dptr, bytesize);
+    return pcuMemAlloc(dptr, bytesize);
+}
+
+CUresult WINAPI wine_cuMemAllocHost(void **pp, size_t bytesize)
+{
+    TRACE("(%p, %lu)\n", pp, (SIZE_T)bytesize);
+    return pcuMemAllocHost(pp, bytesize);
+}
+
+CUresult WINAPI wine_cuMemAllocHost_v2(void **pp, size_t bytesize)
+{
+    TRACE("(%p, %lu)\n", pp, (SIZE_T)bytesize);
+    return pcuMemAllocHost_v2(pp, bytesize);
+}
+
+CUresult WINAPI wine_cuMemAllocManaged(CUdeviceptr *dptr, size_t bytesize, unsigned int flags)
+{
+    TRACE("(%p, %lu, %u)\n", dptr, (SIZE_T)bytesize, flags);
+    return pcuMemAllocManaged(dptr, bytesize, flags);
+}
+
+CUresult WINAPI wine_cuMemAllocPitch(CUdeviceptr *dptr, size_t *pPitch, size_t WidthInBytes,
+                                     size_t Height, unsigned int ElementSizeBytes)
+{
+    TRACE("(%p, %p, %lu, %lu, %u)\n", dptr, pPitch, (SIZE_T)WidthInBytes, (SIZE_T)Height, ElementSizeBytes);
+    return pcuMemAllocPitch(dptr, pPitch, WidthInBytes, Height, ElementSizeBytes);
+}
+
+CUresult WINAPI wine_cuMemAllocPitch_v2(CUdeviceptr *dptr, size_t *pPitch, size_t WidthInBytes,
+                                        size_t Height, unsigned int ElementSizeBytes)
+{
+    TRACE("(%p, %p, %lu, %lu, %u)\n", dptr, pPitch, (SIZE_T)WidthInBytes, (SIZE_T)Height, ElementSizeBytes);
+    return pcuMemAllocPitch_v2(dptr, pPitch, WidthInBytes, Height, ElementSizeBytes);
+}
+
+CUresult WINAPI wine_cuMemAlloc_v2(CUdeviceptr *dptr, unsigned int bytesize)
+{
+    TRACE("(%p, %u)\n", dptr, bytesize);
+    return pcuMemAlloc_v2(dptr, bytesize);
+}
+
+CUresult WINAPI wine_cuMemFree(CUdeviceptr dptr)
+{
+    TRACE("(" DEV_PTR ")\n", dptr);
+    return pcuMemFree(dptr);
+}
+
+CUresult WINAPI wine_cuMemFreeHost(void *p)
+{
+    TRACE("(%p)\n", p);
+    return pcuMemFreeHost(p);
+}
+
+CUresult WINAPI wine_cuMemFree_v2(CUdeviceptr dptr)
+{
+    TRACE("(" DEV_PTR ")\n", dptr);
+    return pcuMemFree_v2(dptr);
+}
+
+CUresult WINAPI wine_cuMemGetAddressRange(CUdeviceptr *pbase, unsigned int *psize, CUdeviceptr dptr)
+{
+    TRACE("(%p, %p, " DEV_PTR ")\n", pbase, psize, dptr);
+    return pcuMemGetAddressRange(pbase, psize, dptr);
+}
+
+CUresult WINAPI wine_cuMemGetAddressRange_v2(CUdeviceptr *pbase, unsigned int *psize, CUdeviceptr dptr)
+{
+    TRACE("(%p, %p, " DEV_PTR ")\n", pbase, psize, dptr);
+    return pcuMemGetAddressRange_v2(pbase, psize, dptr);
+}
+
+CUresult WINAPI wine_cuMemGetInfo(size_t *free, size_t *total)
+{
+    TRACE("(%p, %p)\n", free, total);
+    return pcuMemGetInfo(free, total);
+}
+
+CUresult WINAPI wine_cuMemGetInfo_v2(size_t *free, size_t *total)
+{
+    TRACE("(%p, %p)\n", free, total);
+    return pcuMemGetInfo_v2(free, total);
+}
+
+CUresult WINAPI wine_cuMemHostAlloc(void **pp, size_t bytesize, unsigned int Flags)
+{
+    TRACE("(%p, %lu, %u)\n", pp, (SIZE_T)bytesize, Flags);
+    return pcuMemHostAlloc(pp, bytesize, Flags);
+}
+
+CUresult WINAPI wine_cuMemHostGetDevicePointer(CUdeviceptr *pdptr, void *p, unsigned int Flags)
+{
+    TRACE("(%p, %p, %u)\n", pdptr, p, Flags);
+    return pcuMemHostGetDevicePointer(pdptr, p, Flags);
+}
+
+CUresult WINAPI wine_cuMemHostGetDevicePointer_v2(CUdeviceptr *pdptr, void *p, unsigned int Flags)
+{
+    TRACE("(%p, %p, %u)\n", pdptr, p, Flags);
+    return pcuMemHostGetDevicePointer_v2(pdptr, p, Flags);
+}
+
+CUresult WINAPI wine_cuMemHostGetFlags(unsigned int *pFlags, void *p)
+{
+    TRACE("(%p, %p)\n", pFlags, p);
+    return pcuMemHostGetFlags(pFlags, p);
+}
+
+CUresult WINAPI wine_cuMemHostRegister(void *p, size_t bytesize, unsigned int Flags)
+{
+    TRACE("(%p, %lu, %u)\n", p, (SIZE_T)bytesize, Flags);
+    return pcuMemHostRegister(p, bytesize, Flags);
+}
+
+CUresult WINAPI wine_cuMemHostUnregister(void *p)
+{
+    TRACE("(%p)\n", p);
+    return pcuMemHostUnregister(p);
+}
+
+CUresult WINAPI wine_cuMemcpy(CUdeviceptr dst, CUdeviceptr src, size_t ByteCount)
+{
+    TRACE("(" DEV_PTR ", " DEV_PTR ", %lu)\n", dst, src, (SIZE_T)ByteCount);
+    return pcuMemcpy(dst, src, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpy2D(const CUDA_MEMCPY2D *pCopy)
+{
+    TRACE("(%p)\n", pCopy);
+    return pcuMemcpy2D(pCopy);
+}
+
+CUresult WINAPI wine_cuMemcpy2DAsync(const CUDA_MEMCPY2D *pCopy, CUstream hStream)
+{
+    TRACE("(%p, %p)\n", pCopy, hStream);
+    return pcuMemcpy2DAsync(pCopy, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpy2DAsync_v2(const CUDA_MEMCPY2D *pCopy, CUstream hStream)
+{
+    TRACE("(%p, %p)\n", pCopy, hStream);
+    return pcuMemcpy2DAsync_v2(pCopy, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpy2DUnaligned(const CUDA_MEMCPY2D *pCopy)
+{
+    TRACE("(%p)\n", pCopy);
+    return pcuMemcpy2DUnaligned(pCopy);
+}
+
+CUresult WINAPI wine_cuMemcpy2DUnaligned_v2(const CUDA_MEMCPY2D *pCopy)
+{
+    TRACE("(%p)\n", pCopy);
+    return pcuMemcpy2DUnaligned_v2(pCopy);
+}
+
+CUresult WINAPI wine_cuMemcpy2D_v2(const CUDA_MEMCPY2D *pCopy)
+{
+    TRACE("(%p)\n", pCopy);
+    return pcuMemcpy2D_v2(pCopy);
+}
+
+CUresult WINAPI wine_cuMemcpy3D(const CUDA_MEMCPY3D *pCopy)
+{
+    TRACE("(%p)\n", pCopy);
+    return pcuMemcpy3D(pCopy);
+}
+
+CUresult WINAPI wine_cuMemcpy3DAsync(const CUDA_MEMCPY3D *pCopy, CUstream hStream)
+{
+    TRACE("(%p, %p)\n", pCopy, hStream);
+    return pcuMemcpy3DAsync(pCopy, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpy3DAsync_v2(const CUDA_MEMCPY3D *pCopy, CUstream hStream)
+{
+    TRACE("(%p, %p)\n", pCopy, hStream);
+    return pcuMemcpy3DAsync_v2(pCopy, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpy3DPeer(const CUDA_MEMCPY3D_PEER *pCopy)
+{
+    TRACE("(%p)\n", pCopy);
+    return pcuMemcpy3DPeer(pCopy);
+}
+
+CUresult WINAPI wine_cuMemcpy3DPeerAsync(const CUDA_MEMCPY3D_PEER *pCopy, CUstream hStream)
+{
+    TRACE("(%p, %p)\n", pCopy, hStream);
+    return pcuMemcpy3DPeerAsync(pCopy, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpy3D_v2(const CUDA_MEMCPY3D *pCopy)
+{
+    TRACE("(%p)\n", pCopy);
+    return pcuMemcpy3D_v2(pCopy);
+}
+
+CUresult WINAPI wine_cuMemcpyAsync(CUdeviceptr dst, CUdeviceptr src, size_t ByteCount, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", " DEV_PTR ", %lu, %p)\n", dst, src, (SIZE_T)ByteCount, hStream);
+    return pcuMemcpyAsync(dst, src, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyAtoA(CUarray dstArray, size_t dstOffset, CUarray srcArray,
+                                  size_t srcOffset, size_t ByteCount)
+{
+    TRACE("(%p, %lu, %p, %lu, %lu)\n", dstArray, (SIZE_T)dstOffset, srcArray, (SIZE_T)srcOffset, (SIZE_T)ByteCount);
+    return pcuMemcpyAtoA(dstArray, dstOffset, srcArray, srcOffset, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyAtoA_v2(CUarray dstArray, size_t dstOffset, CUarray srcArray,
+                                     size_t srcOffset, size_t ByteCount)
+{
+    TRACE("(%p, %lu, %p, %lu, %lu)\n", dstArray, (SIZE_T)dstOffset, srcArray, (SIZE_T)srcOffset, (SIZE_T)ByteCount);
+    return pcuMemcpyAtoA_v2(dstArray, dstOffset, srcArray, srcOffset, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyAtoD(CUdeviceptr dstDevice, CUarray srcArray, size_t srcOffset, size_t ByteCount)
+{
+    TRACE("(" DEV_PTR ", %p, %lu, %lu)\n", dstDevice, srcArray, (SIZE_T)srcOffset, (SIZE_T)ByteCount);
+    return pcuMemcpyAtoD(dstDevice, srcArray, srcOffset, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyAtoD_v2(CUdeviceptr dstDevice, CUarray srcArray, size_t srcOffset, size_t ByteCount)
+{
+    TRACE("(" DEV_PTR ", %p, %lu, %lu)\n", dstDevice, srcArray, (SIZE_T)srcOffset, (SIZE_T)ByteCount);
+    return pcuMemcpyAtoD_v2(dstDevice, srcArray, srcOffset, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyAtoH(void *dstHost, CUarray srcArray, size_t srcOffset, size_t ByteCount)
+{
+    TRACE("(%p, %p, %lu, %lu)\n", dstHost, srcArray, (SIZE_T)srcOffset, (SIZE_T)ByteCount);
+    return pcuMemcpyAtoH(dstHost, srcArray, srcOffset, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyAtoHAsync(void *dstHost, CUarray srcArray, size_t srcOffset, size_t ByteCount, CUstream hStream)
+{
+    TRACE("(%p, %p, %lu, %lu, %p)\n", dstHost, srcArray, (SIZE_T)srcOffset, (SIZE_T)ByteCount, hStream);
+    return pcuMemcpyAtoHAsync(dstHost, srcArray, srcOffset, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyAtoHAsync_v2(void *dstHost, CUarray srcArray, size_t srcOffset, size_t ByteCount, CUstream hStream)
+{
+    TRACE("(%p, %p, %lu, %lu, %p)\n", dstHost, srcArray, (SIZE_T)srcOffset, (SIZE_T)ByteCount, hStream);
+    return pcuMemcpyAtoHAsync_v2(dstHost, srcArray, srcOffset, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyAtoH_v2(void *dstHost, CUarray srcArray, size_t srcOffset, size_t ByteCount)
+{
+    TRACE("(%p, %p, %lu, %lu)\n", dstHost, srcArray, (SIZE_T)srcOffset, (SIZE_T)ByteCount);
+    return pcuMemcpyAtoH_v2(dstHost, srcArray, srcOffset, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoA(CUarray dstArray, size_t dstOffset, CUdeviceptr srcDevice, size_t ByteCount)
+{
+    TRACE("(%p, %lu, " DEV_PTR ", %lu)\n", dstArray, (SIZE_T)dstOffset, srcDevice, (SIZE_T)ByteCount);
+    return pcuMemcpyDtoA(dstArray, dstOffset, srcDevice, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoA_v2(CUarray dstArray, size_t dstOffset, CUdeviceptr srcDevice, size_t ByteCount)
+{
+    TRACE("(%p, %lu, " DEV_PTR ", %lu)\n", dstArray, (SIZE_T)dstOffset, srcDevice, (SIZE_T)ByteCount);
+    return pcuMemcpyDtoA_v2(dstArray, dstOffset, srcDevice, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoD(CUdeviceptr dstDevice, CUdeviceptr srcDevice, unsigned int ByteCount)
+{
+    TRACE("(" DEV_PTR ", " DEV_PTR ", %u)\n", dstDevice, srcDevice, ByteCount);
+    return pcuMemcpyDtoD(dstDevice, srcDevice, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoDAsync(CUdeviceptr dstDevice, CUdeviceptr srcDevice,
+                                       unsigned int ByteCount, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", " DEV_PTR ", %u, %p)\n", dstDevice, srcDevice, ByteCount, hStream);
+    return pcuMemcpyDtoDAsync(dstDevice, srcDevice, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoDAsync_v2(CUdeviceptr dstDevice, CUdeviceptr srcDevice,
+                                          unsigned int ByteCount, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", " DEV_PTR ", %u, %p)\n", dstDevice, srcDevice, ByteCount, hStream);
+    return pcuMemcpyDtoDAsync_v2(dstDevice, srcDevice, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoD_v2(CUdeviceptr dstDevice, CUdeviceptr srcDevice, unsigned int ByteCount)
+{
+    TRACE("(" DEV_PTR ", " DEV_PTR ", %u)\n", dstDevice, srcDevice, ByteCount);
+    return pcuMemcpyDtoD_v2(dstDevice, srcDevice, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoH(void *dstHost, CUdeviceptr srcDevice, unsigned int ByteCount)
+{
+    TRACE("(%p, " DEV_PTR ", %u)\n", dstHost, srcDevice, ByteCount);
+    return pcuMemcpyDtoH(dstHost, srcDevice, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoHAsync(void *dstHost, CUdeviceptr srcDevice, unsigned int ByteCount, CUstream hStream)
+{
+    TRACE("(%p, " DEV_PTR ", %u, %p)\n", dstHost, srcDevice, ByteCount, hStream);
+    return pcuMemcpyDtoHAsync(dstHost, srcDevice, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoHAsync_v2(void *dstHost, CUdeviceptr srcDevice, unsigned int ByteCount, CUstream hStream)
+{
+    TRACE("(%p, " DEV_PTR ", %u, %p)\n", dstHost, srcDevice, ByteCount, hStream);
+    return pcuMemcpyDtoHAsync_v2(dstHost, srcDevice, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyDtoH_v2(void *dstHost, CUdeviceptr srcDevice, unsigned int ByteCount)
+{
+    TRACE("(%p, " DEV_PTR ", %u)\n", dstHost, srcDevice, ByteCount);
+    return pcuMemcpyDtoH_v2(dstHost, srcDevice, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyHtoA(CUarray dstArray, size_t dstOffset, const void *srcHost, size_t ByteCount)
+{
+    TRACE("(%p, %lu, %p, %lu)\n", dstArray, (SIZE_T)dstOffset, srcHost, (SIZE_T)ByteCount);
+    return pcuMemcpyHtoA(dstArray, dstOffset, srcHost, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyHtoAAsync(CUarray dstArray, size_t dstOffset, const void *srcHost,
+                                       size_t ByteCount, CUstream hStream)
+{
+    TRACE("(%p, %lu, %p, %lu, %p)\n", dstArray, (SIZE_T)dstOffset, srcHost, (SIZE_T)ByteCount, hStream);
+    return pcuMemcpyHtoAAsync(dstArray, dstOffset, srcHost, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyHtoAAsync_v2(CUarray dstArray, size_t dstOffset, const void *srcHost,
+                                          size_t ByteCount, CUstream hStream)
+{
+    TRACE("(%p, %lu, %p, %lu, %p)\n", dstArray, (SIZE_T)dstOffset, srcHost, (SIZE_T)ByteCount, hStream);
+    return pcuMemcpyHtoAAsync_v2(dstArray, dstOffset, srcHost, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyHtoA_v2(CUarray dstArray, size_t dstOffset, const void *srcHost, size_t ByteCount)
+{
+    TRACE("(%p, %lu, %p, %lu)\n", dstArray, (SIZE_T)dstOffset, srcHost, (SIZE_T)ByteCount);
+    return pcuMemcpyHtoA_v2(dstArray, dstOffset, srcHost, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyHtoD(CUdeviceptr dstDevice, const void *srcHost, size_t ByteCount)
+{
+    TRACE("(" DEV_PTR ", %p, %lu)\n", dstDevice, srcHost, (SIZE_T)ByteCount);
+    return pcuMemcpyHtoD(dstDevice, srcHost, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyHtoDAsync(CUdeviceptr dstDevice, const void *srcHost, size_t ByteCount, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", %p, %lu, %p)\n", dstDevice, srcHost, (SIZE_T)ByteCount, hStream);
+    return pcuMemcpyHtoDAsync(dstDevice, srcHost, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyHtoDAsync_v2(CUdeviceptr dstDevice, const void *srcHost, size_t ByteCount, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", %p, %lu, %p)\n", dstDevice, srcHost, (SIZE_T)ByteCount, hStream);
+    return pcuMemcpyHtoDAsync_v2(dstDevice, srcHost, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemcpyHtoD_v2(CUdeviceptr dstDevice, const void *srcHost, size_t ByteCount)
+{
+    TRACE("(" DEV_PTR ", %p, %lu)\n", dstDevice, srcHost, (SIZE_T)ByteCount);
+    return pcuMemcpyHtoD_v2(dstDevice, srcHost, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyPeer(CUdeviceptr dstDevice, CUcontext dstContext, CUdeviceptr srcDevice,
+                                  CUcontext srcContext, size_t ByteCount)
+{
+    TRACE("(" DEV_PTR ", %p, " DEV_PTR ", %p, %lu)\n", dstDevice, dstContext, srcDevice, srcContext, (SIZE_T)ByteCount);
+    return pcuMemcpyPeer(dstDevice, dstContext, srcDevice, srcContext, ByteCount);
+}
+
+CUresult WINAPI wine_cuMemcpyPeerAsync(CUdeviceptr dstDevice, CUcontext dstContext, CUdeviceptr srcDevice,
+                                       CUcontext srcContext, size_t ByteCount, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", %p, " DEV_PTR ", %p, %lu, %p)\n", dstDevice, dstContext, srcDevice, srcContext, (SIZE_T)ByteCount, hStream);
+    return pcuMemcpyPeerAsync(dstDevice, dstContext, srcDevice, srcContext, ByteCount, hStream);
+}
+
+CUresult WINAPI wine_cuMemsetD16(CUdeviceptr dstDevice, unsigned short us, size_t N)
+{
+    TRACE("(" DEV_PTR ", %u, %lu)\n", dstDevice, us, (SIZE_T)N);
+    return pcuMemsetD16(dstDevice, us, N);
+}
+
+CUresult WINAPI wine_cuMemsetD16Async(CUdeviceptr dstDevice, unsigned short us, size_t N, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", %u, %lu, %p)\n", dstDevice, us, (SIZE_T)N, hStream);
+    return pcuMemsetD16Async(dstDevice, us, N, hStream);
+}
+
+CUresult WINAPI wine_cuMemsetD16_v2(CUdeviceptr dstDevice, unsigned short us, size_t N)
+{
+    TRACE("(" DEV_PTR ", %u, %lu)\n", dstDevice, us, (SIZE_T)N);
+    return pcuMemsetD16_v2(dstDevice, us, N);
+}
+
+CUresult WINAPI wine_cuMemsetD2D16(CUdeviceptr dstDevice, size_t dstPitch, unsigned short us, size_t Width, size_t Height)
+{
+    TRACE("(" DEV_PTR ", %lu, %u, %lu, %lu)\n", dstDevice, (SIZE_T)dstPitch, us, (SIZE_T)Width, (SIZE_T)Height);
+    return pcuMemsetD2D16(dstDevice, dstPitch, us, Width, Height);
+}
+
+CUresult WINAPI wine_cuMemsetD2D16Async(CUdeviceptr dstDevice, size_t dstPitch, unsigned short us,
+                                        size_t Width, size_t Height, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", %lu, %u, %lu, %lu, %p)\n", dstDevice, (SIZE_T)dstPitch, us, (SIZE_T)Width, (SIZE_T)Height, hStream);
+    return pcuMemsetD2D16Async(dstDevice, dstPitch, us, Width, Height, hStream);
+}
+
+CUresult WINAPI wine_cuMemsetD2D16_v2(CUdeviceptr dstDevice, size_t dstPitch, unsigned short us, size_t Width, size_t Height)
+{
+    TRACE("(" DEV_PTR ", %lu, %u, %lu, %lu)\n", dstDevice, (SIZE_T)dstPitch, us, (SIZE_T)Width, (SIZE_T)Height);
+    return pcuMemsetD2D16_v2(dstDevice, dstPitch, us, Width, Height);
+}
+
+CUresult WINAPI wine_cuMemsetD2D32(CUdeviceptr dstDevice, size_t dstPitch, unsigned int ui, size_t Width, size_t Height)
+{
+    TRACE("(" DEV_PTR ", %lu, %u, %lu, %lu)\n", dstDevice, (SIZE_T)dstPitch, ui, (SIZE_T)Width, (SIZE_T)Height);
+    return pcuMemsetD2D32(dstDevice, dstPitch, ui, Width, Height);
+}
+
+CUresult WINAPI wine_cuMemsetD2D32Async(CUdeviceptr dstDevice, size_t dstPitch, unsigned int ui,
+                                        size_t Width, size_t Height, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", %lu, %u, %lu, %lu, %p)\n", dstDevice, (SIZE_T)dstPitch, ui, (SIZE_T)Width, (SIZE_T)Height, hStream);
+    return pcuMemsetD2D32Async(dstDevice, dstPitch, ui, Width, Height, hStream);
+}
+
+CUresult WINAPI wine_cuMemsetD2D32_v2(CUdeviceptr dstDevice, size_t dstPitch, unsigned int ui, size_t Width, size_t Height)
+{
+    TRACE("(" DEV_PTR ", %lu, %u, %lu, %lu)\n", dstDevice, (SIZE_T)dstPitch, ui, (SIZE_T)Width, (SIZE_T)Height);
+    return pcuMemsetD2D32_v2(dstDevice, dstPitch, ui, Width, Height);
+}
+
+CUresult WINAPI wine_cuMemsetD2D8(CUdeviceptr dstDevice, unsigned int dstPitch, unsigned char uc,
+                                  unsigned int Width, unsigned int Height)
+{
+    TRACE("(" DEV_PTR ", %u, %x, %u, %u)\n", dstDevice, dstPitch, uc, Width, Height);
+    return pcuMemsetD2D8(dstDevice, dstPitch, uc, Width, Height);
+}
+
+CUresult WINAPI wine_cuMemsetD2D8Async(CUdeviceptr dstDevice, size_t dstPitch, unsigned char uc,
+                                       size_t Width, size_t Height, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", %lu, %u, %lu, %lu, %p)\n", dstDevice, (SIZE_T)dstPitch, uc, (SIZE_T)Width, (SIZE_T)Height, hStream);
+    return pcuMemsetD2D8Async(dstDevice, dstPitch, uc, Width, Height, hStream);
+}
+
+CUresult WINAPI wine_cuMemsetD2D8_v2(CUdeviceptr dstDevice, unsigned int dstPitch, unsigned char uc,
+                                     unsigned int Width, unsigned int Height)
+{
+    TRACE("(" DEV_PTR ", %u, %x, %u, %u)\n", dstDevice, dstPitch, uc, Width, Height);
+    return pcuMemsetD2D8_v2(dstDevice, dstPitch, uc, Width, Height);
+}
+
+CUresult WINAPI wine_cuMemsetD32(CUdeviceptr dstDevice, unsigned int ui, size_t N)
+{
+    TRACE("(" DEV_PTR ", %u, %lu)\n", dstDevice, ui, (SIZE_T)N);
+    return pcuMemsetD32(dstDevice, ui, N);
+}
+
+CUresult WINAPI wine_cuMemsetD32Async(CUdeviceptr dstDevice, unsigned int ui, size_t N, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", %u, %lu, %p)\n", dstDevice, ui, (SIZE_T)N, hStream);
+    return pcuMemsetD32Async(dstDevice, ui, N, hStream);
+}
+
+CUresult WINAPI wine_cuMemsetD32_v2(CUdeviceptr dstDevice, unsigned int ui, size_t N)
+{
+    TRACE("(" DEV_PTR ", %u, %lu)\n", dstDevice, ui, (SIZE_T)N);
+    return pcuMemsetD32_v2(dstDevice, ui, N);
+}
+
+CUresult WINAPI wine_cuMemsetD8(CUdeviceptr dstDevice, unsigned char uc, unsigned int N)
+{
+    TRACE("(" DEV_PTR ", %x, %u)\n", dstDevice, uc, N);
+    return pcuMemsetD8(dstDevice, uc, N);
+}
+
+CUresult WINAPI wine_cuMemsetD8Async(CUdeviceptr dstDevice, unsigned char uc, size_t N, CUstream hStream)
+{
+    TRACE("(" DEV_PTR ", %x, %lu, %p)\n", dstDevice, uc, (SIZE_T)N, hStream);
+    return pcuMemsetD8Async(dstDevice, uc, N, hStream);
+}
+
+CUresult WINAPI wine_cuMemsetD8_v2(CUdeviceptr dstDevice, unsigned char uc, unsigned int N)
+{
+    TRACE("(" DEV_PTR ", %x, %u)\n", dstDevice, uc, N);
+    return pcuMemsetD8_v2(dstDevice, uc, N);
+}
+
+CUresult WINAPI wine_cuMipmappedArrayCreate(CUmipmappedArray *pHandle, const CUDA_ARRAY3D_DESCRIPTOR *pMipmappedArrayDesc,
+                                            unsigned int numMipmapLevels)
+{
+    TRACE("(%p, %p, %u)\n", pHandle, pMipmappedArrayDesc, numMipmapLevels);
+    return pcuMipmappedArrayCreate(pHandle, pMipmappedArrayDesc, numMipmapLevels);
+}
+
+CUresult WINAPI wine_cuMipmappedArrayDestroy(CUmipmappedArray hMipmappedArray)
+{
+    TRACE("(%p)\n", hMipmappedArray);
+    return pcuMipmappedArrayDestroy(hMipmappedArray);
+}
+
+CUresult WINAPI wine_cuMipmappedArrayGetLevel(CUarray *pLevelArray, CUmipmappedArray hMipmappedArray, unsigned int level)
+{
+    TRACE("(%p, %p, %u)\n", pLevelArray, hMipmappedArray, level);
+    return pcuMipmappedArrayGetLevel(pLevelArray, hMipmappedArray, level);
+}
+
+CUresult WINAPI wine_cuModuleGetFunction(CUfunction *hfunc, CUmodule hmod, const char *name)
+{
+    TRACE("(%p, %p, %s)\n", hfunc, hmod, name);
+    return pcuModuleGetFunction(hfunc, hmod, name);
+}
+
+CUresult WINAPI wine_cuModuleGetGlobal(CUdeviceptr *dptr, size_t *bytes, CUmodule hmod, const char *name)
+{
+    TRACE("(%p, %p, %p, %s)\n", dptr, bytes, hmod, name);
+    return pcuModuleGetGlobal(dptr, bytes, hmod, name);
+}
+
+CUresult WINAPI wine_cuModuleGetGlobal_v2(CUdeviceptr *dptr, size_t *bytes, CUmodule hmod, const char *name)
+{
+    TRACE("(%p, %p, %p, %s)\n", dptr, bytes, hmod, name);
+    return pcuModuleGetGlobal_v2(dptr, bytes, hmod, name);
+}
+
+CUresult WINAPI wine_cuModuleGetSurfRef(CUsurfref *pSurfRef, CUmodule hmod, const char *name)
+{
+    TRACE("(%p, %p, %s)\n", pSurfRef, hmod, name);
+    return pcuModuleGetSurfRef(pSurfRef, hmod, name);
+}
+
+CUresult WINAPI wine_cuModuleGetTexRef(CUtexref *pTexRef, CUmodule hmod, const char *name)
+{
+    TRACE("(%p, %p, %s)\n", pTexRef, hmod, name);
+    return pcuModuleGetTexRef(pTexRef, hmod, name);
+}
+
+CUresult WINAPI wine_cuModuleLoadData(CUmodule *module, const void *image)
+{
+    TRACE("(%p, %p)\n", module, image);
+    return pcuModuleLoadData(module, image);
+}
+
+CUresult WINAPI wine_cuModuleLoadDataEx(CUmodule *module, const void *image, unsigned int numOptions,
+                                        CUjit_option *options, void **optionValues)
+{
+    TRACE("(%p, %p, %u, %p, %p)\n", module, image, numOptions, options, optionValues);
+    return pcuModuleLoadDataEx(module, image, numOptions, options, optionValues);
+}
+
+CUresult WINAPI wine_cuModuleLoadFatBinary(CUmodule *module, const void *fatCubin)
+{
+    TRACE("(%p, %p)\n", module, fatCubin);
+    return pcuModuleLoadFatBinary(module, fatCubin);
+}
+
+CUresult WINAPI wine_cuModuleUnload(CUmodule hmod)
+{
+    TRACE("(%p)\n", hmod);
+    return pcuModuleUnload(hmod);
+}
+
+CUresult WINAPI wine_cuParamSetSize(CUfunction hfunc, unsigned int numbytes)
+{
+    TRACE("(%p, %u)\n", hfunc, numbytes);
+    return pcuParamSetSize(hfunc, numbytes);
+}
+
+CUresult WINAPI wine_cuParamSetTexRef(CUfunction hfunc, int texunit, CUtexref hTexRef)
+{
+    TRACE("(%p, %d, %p)\n", hfunc, texunit, hTexRef);
+    return pcuParamSetTexRef(hfunc, texunit, hTexRef);
+}
+
+CUresult WINAPI wine_cuParamSetf(CUfunction hfunc, int offset, float value)
+{
+    TRACE("(%p, %d, %f)\n", hfunc, offset, value);
+    return pcuParamSetf(hfunc, offset, value);
+}
+
+CUresult WINAPI wine_cuParamSeti(CUfunction hfunc, int offset, unsigned int value)
+{
+    TRACE("(%p, %d, %u)\n", hfunc, offset, value);
+    return pcuParamSeti(hfunc, offset, value);
+}
+
+CUresult WINAPI wine_cuParamSetv(CUfunction hfunc, int offset, void *ptr, unsigned int numbytes)
+{
+    TRACE("(%p, %d, %p, %u)\n", hfunc, offset, ptr, numbytes);
+    return pcuParamSetv(hfunc, offset, ptr, numbytes);
+}
+
+CUresult WINAPI wine_cuPointerGetAttribute(void *data, CUpointer_attribute attribute, CUdeviceptr ptr)
+{
+    TRACE("(%p, %d, " DEV_PTR ")\n", data, attribute, ptr);
+    return pcuPointerGetAttribute(data, attribute, ptr);
+}
+
+CUresult WINAPI wine_cuPointerSetAttribute(const void *value, CUpointer_attribute attribute, CUdeviceptr ptr)
+{
+    TRACE("(%p, %d, " DEV_PTR ")\n", value, attribute, ptr);
+    return pcuPointerSetAttribute(value, attribute, ptr);
+}
+
+struct stream_callback
+{
+    void (WINAPI *callback)(CUstream hStream, CUresult status, void *userData);
+    void *userData;
+};
+
+static void stream_callback_wrapper(CUstream hStream, CUresult status, void *userData)
+{
+    struct stream_callback *wrapper = userData;
+    TRACE("(%p, %d, %p)\n", hStream, status, userData);
+
+    TRACE("calling stream callback %p(%p, %d, %p)\n", wrapper->callback, hStream, status, wrapper->userData);
+    wrapper->callback(hStream, status, wrapper->userData);
+    TRACE("stream callback %p returned\n", wrapper->callback);
+
+    HeapFree( GetProcessHeap(), 0, wrapper );
+}
+
+CUresult WINAPI wine_cuStreamAddCallback(CUstream hStream, void *callback, void *userData, unsigned int flags)
+{
+    struct stream_callback *wrapper;
+    CUresult ret;
+
+    TRACE("(%p, %p, %p, %u)\n", hStream, callback, userData, flags);
+
+    wrapper = HeapAlloc( GetProcessHeap(), 0, sizeof(*wrapper) );
+    if (!wrapper)
+        return CUDA_ERROR_OUT_OF_MEMORY;
+
+    wrapper->callback = callback;
+    wrapper->userData = userData;
+
+    ret = pcuStreamAddCallback(hStream, stream_callback_wrapper, wrapper, flags);
+    if (ret) HeapFree( GetProcessHeap(), 0, wrapper );
+    return ret;
+}
+
+CUresult WINAPI wine_cuStreamAttachMemAsync(CUstream hStream, CUdeviceptr dptr, size_t length, unsigned int flags)
+{
+    TRACE("(%p, " DEV_PTR ", %lu, %u)\n", hStream, dptr, (SIZE_T)length, flags);
+    return pcuStreamAttachMemAsync(hStream, dptr, length, flags);
+}
+
+CUresult WINAPI wine_cuStreamCreate(CUstream *phStream, unsigned int Flags)
+{
+    TRACE("(%p, %u)\n", phStream, Flags);
+    return pcuStreamCreate(phStream, Flags);
+}
+
+CUresult WINAPI wine_cuStreamCreateWithPriority(CUstream *phStream, unsigned int flags, int priority)
+{
+    TRACE("(%p, %u, %d)\n", phStream, flags, priority);
+    return pcuStreamCreateWithPriority(phStream, flags, priority);
+}
+
+CUresult WINAPI wine_cuStreamDestroy(CUstream hStream)
+{
+    TRACE("(%p)\n", hStream);
+    return pcuStreamDestroy(hStream);
+}
+
+CUresult WINAPI wine_cuStreamDestroy_v2(CUstream hStream)
+{
+    TRACE("(%p)\n", hStream);
+    return pcuStreamDestroy_v2(hStream);
+}
+
+CUresult WINAPI wine_cuStreamGetFlags(CUstream hStream, unsigned int *flags)
+{
+    TRACE("(%p, %p)\n", hStream, flags);
+    return pcuStreamGetFlags(hStream, flags);
+}
+
+CUresult WINAPI wine_cuStreamGetPriority(CUstream hStream, int *priority)
+{
+    TRACE("(%p, %p)\n", hStream, priority);
+    return pcuStreamGetPriority(hStream, priority);
+}
+
+CUresult WINAPI wine_cuStreamQuery(CUstream hStream)
+{
+    TRACE("(%p)\n", hStream);
+    return pcuStreamQuery(hStream);
+}
+
+CUresult WINAPI wine_cuStreamSynchronize(CUstream hStream)
+{
+    TRACE("(%p)\n", hStream);
+    return pcuStreamSynchronize(hStream);
+}
+
+CUresult WINAPI wine_cuStreamWaitEvent(CUstream hStream, CUevent hEvent, unsigned int Flags)
+{
+    TRACE("(%p, %p, %u)\n", hStream, hEvent, Flags);
+    return pcuStreamWaitEvent(hStream, hEvent, Flags);
+}
+
+CUresult WINAPI wine_cuSurfObjectCreate(CUsurfObject *pSurfObject, const CUDA_RESOURCE_DESC *pResDesc)
+{
+    TRACE("(%p, %p)\n", pSurfObject, pResDesc);
+    return pcuSurfObjectCreate(pSurfObject, pResDesc);
+}
+
+CUresult WINAPI wine_cuSurfObjectDestroy(CUsurfObject surfObject)
+{
+    TRACE("(%llu)\n", surfObject);
+    return pcuSurfObjectDestroy(surfObject);
+}
+
+CUresult WINAPI wine_cuSurfObjectGetResourceDesc(CUDA_RESOURCE_DESC *pResDesc, CUsurfObject surfObject)
+{
+    TRACE("(%p, %llu)\n", pResDesc, surfObject);
+    return pcuSurfObjectGetResourceDesc(pResDesc, surfObject);
+}
+
+CUresult WINAPI wine_cuSurfRefGetArray(CUarray *phArray, CUsurfref hSurfRef)
+{
+    TRACE("(%p, %p)\n", phArray, hSurfRef);
+    return pcuSurfRefGetArray(phArray, hSurfRef);
+}
+
+CUresult WINAPI wine_cuSurfRefSetArray(CUsurfref hSurfRef, CUarray hArray, unsigned int Flags)
+{
+    TRACE("(%p, %p, %u)\n", hSurfRef, hArray, Flags);
+    return pcuSurfRefSetArray(hSurfRef, hArray, Flags);
+}
+
+CUresult WINAPI wine_cuTexObjectCreate(CUtexObject *pTexObject, const CUDA_RESOURCE_DESC *pResDesc,
+                                       const CUDA_TEXTURE_DESC *pTexDesc, const CUDA_RESOURCE_VIEW_DESC *pResViewDesc)
+{
+    TRACE("(%p, %p, %p, %p)\n", pTexObject, pResDesc, pTexDesc, pResViewDesc);
+    return pcuTexObjectCreate(pTexObject, pResDesc, pTexDesc, pResViewDesc);
+}
+
+CUresult WINAPI wine_cuTexObjectDestroy(CUtexObject texObject)
+{
+    TRACE("(%llu)\n", texObject);
+    return pcuTexObjectDestroy(texObject);
+}
+
+CUresult WINAPI wine_cuTexObjectGetResourceDesc(CUDA_RESOURCE_DESC *pResDesc, CUtexObject texObject)
+{
+    TRACE("(%p, %llu)\n", pResDesc, texObject);
+    return pcuTexObjectGetResourceDesc(pResDesc, texObject);
+}
+
+CUresult WINAPI wine_cuTexObjectGetResourceViewDesc(CUDA_RESOURCE_VIEW_DESC *pResViewDesc, CUtexObject texObject)
+{
+    TRACE("(%p, %llu)\n", pResViewDesc, texObject);
+    return pcuTexObjectGetResourceViewDesc(pResViewDesc, texObject);
+}
+
+CUresult WINAPI wine_cuTexObjectGetTextureDesc(CUDA_TEXTURE_DESC *pTexDesc, CUtexObject texObject)
+{
+    TRACE("(%p, %llu)\n", pTexDesc, texObject);
+    return pcuTexObjectGetTextureDesc(pTexDesc, texObject);
+}
+
+CUresult WINAPI wine_cuTexRefCreate(CUtexref *pTexRef)
+{
+    TRACE("(%p)\n", pTexRef);
+    return pcuTexRefCreate(pTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefDestroy(CUtexref hTexRef)
+{
+    TRACE("(%p)\n", hTexRef);
+    return pcuTexRefDestroy(hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetAddress(CUdeviceptr *pdptr, CUtexref hTexRef)
+{
+    TRACE("(%p, %p)\n", pdptr, hTexRef);
+    return pcuTexRefGetAddress(pdptr, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetAddressMode(CUaddress_mode *pam, CUtexref hTexRef, int dim)
+{
+    TRACE("(%p, %p, %d)\n", pam, hTexRef, dim);
+    return pcuTexRefGetAddressMode(pam, hTexRef, dim);
+}
+
+CUresult WINAPI wine_cuTexRefGetAddress_v2(CUdeviceptr *pdptr, CUtexref hTexRef)
+{
+    TRACE("(%p, %p)\n", pdptr, hTexRef);
+    return pcuTexRefGetAddress_v2(pdptr, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetArray(CUarray *phArray, CUtexref hTexRef)
+{
+    TRACE("(%p, %p)\n", phArray, hTexRef);
+    return pcuTexRefGetArray(phArray, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetFilterMode(CUfilter_mode *pfm, CUtexref hTexRef)
+{
+    TRACE("(%p, %p)\n", pfm, hTexRef);
+    return pcuTexRefGetFilterMode(pfm, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetFlags(unsigned int *pFlags, CUtexref hTexRef)
+{
+    TRACE("(%p, %p)\n", pFlags, hTexRef);
+    return pcuTexRefGetFlags(pFlags, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetFormat(CUarray_format *pFormat, int *pNumChannels, CUtexref hTexRef)
+{
+    TRACE("(%p, %p, %p)\n", pFormat, pNumChannels, hTexRef);
+    return pcuTexRefGetFormat(pFormat, pNumChannels, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetMaxAnisotropy(int *pmaxAniso, CUtexref hTexRef)
+{
+    TRACE("(%p, %p)\n", pmaxAniso, hTexRef);
+    return pcuTexRefGetMaxAnisotropy(pmaxAniso, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetMipmapFilterMode(CUfilter_mode *pfm, CUtexref hTexRef)
+{
+    TRACE("(%p, %p)\n", pfm, hTexRef);
+    return pcuTexRefGetMipmapFilterMode(pfm, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetMipmapLevelBias(float *pbias, CUtexref hTexRef)
+{
+    TRACE("(%p, %p)\n", pbias, hTexRef);
+    return pcuTexRefGetMipmapLevelBias(pbias, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetMipmapLevelClamp(float *pminMipmapLevelClamp, float *pmaxMipmapLevelClamp, CUtexref hTexRef)
+{
+    TRACE("(%p, %p, %p)\n", pminMipmapLevelClamp, pmaxMipmapLevelClamp, hTexRef);
+    return pcuTexRefGetMipmapLevelClamp(pminMipmapLevelClamp, pmaxMipmapLevelClamp, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefGetMipmappedArray(CUmipmappedArray *phMipmappedArray, CUtexref hTexRef)
+{
+    TRACE("(%p, %p)\n", phMipmappedArray, hTexRef);
+    return pcuTexRefGetMipmappedArray(phMipmappedArray, hTexRef);
+}
+
+CUresult WINAPI wine_cuTexRefSetAddress(size_t *ByteOffset, CUtexref hTexRef, CUdeviceptr dptr, size_t bytes)
+{
+    TRACE("(%p, %p, " DEV_PTR ", %lu)\n", ByteOffset, hTexRef, dptr, (SIZE_T)bytes);
+    return pcuTexRefSetAddress(ByteOffset, hTexRef, dptr, bytes);
+}
+
+CUresult WINAPI wine_cuTexRefSetAddress2D(CUtexref hTexRef, const CUDA_ARRAY_DESCRIPTOR *desc,
+                                          CUdeviceptr dptr, unsigned int Pitch)
+{
+    TRACE("(%p, %p, " DEV_PTR ", %u)\n", hTexRef, desc, dptr, Pitch);
+    return pcuTexRefSetAddress2D(hTexRef, desc, dptr, Pitch);
+}
+
+CUresult WINAPI wine_cuTexRefSetAddress2D_v2(CUtexref hTexRef, const CUDA_ARRAY_DESCRIPTOR *desc,
+                                             CUdeviceptr dptr, unsigned int Pitch)
+{
+    TRACE("(%p, %p, " DEV_PTR ", %u)\n", hTexRef, desc, dptr, Pitch);
+    return pcuTexRefSetAddress2D_v2(hTexRef, desc, dptr, Pitch);
+}
+
+CUresult WINAPI wine_cuTexRefSetAddress2D_v3(CUtexref hTexRef, const CUDA_ARRAY_DESCRIPTOR *desc,
+                                             CUdeviceptr dptr, unsigned int Pitch)
+{
+    TRACE("(%p, %p, " DEV_PTR ", %u)\n", hTexRef, desc, dptr, Pitch);
+    return pcuTexRefSetAddress2D_v3(hTexRef, desc, dptr, Pitch);
+}
+
+CUresult WINAPI wine_cuTexRefSetAddressMode(CUtexref hTexRef, int dim, CUaddress_mode am)
+{
+    TRACE("(%p, %d, %u)\n", hTexRef, dim, am);
+    return pcuTexRefSetAddressMode(hTexRef, dim, am);
+}
+
+CUresult WINAPI wine_cuTexRefSetAddress_v2(size_t *ByteOffset, CUtexref hTexRef, CUdeviceptr dptr, size_t bytes)
+{
+    TRACE("(%p, %p, " DEV_PTR ", %lu)\n", ByteOffset, hTexRef, dptr, (SIZE_T)bytes);
+    return pcuTexRefSetAddress_v2(ByteOffset, hTexRef, dptr, bytes);
+}
+
+CUresult WINAPI wine_cuTexRefSetArray(CUtexref hTexRef, CUarray hArray, unsigned int Flags)
+{
+    TRACE("(%p, %p, %u)\n", hTexRef, hArray, Flags);
+    return pcuTexRefSetArray(hTexRef, hArray, Flags);
+}
+
+CUresult WINAPI wine_cuTexRefSetFilterMode(CUtexref hTexRef, CUfilter_mode fm)
+{
+    TRACE("(%p, %u)\n", hTexRef, fm);
+    return pcuTexRefSetFilterMode(hTexRef, fm);
+}
+
+CUresult WINAPI wine_cuTexRefSetFlags(CUtexref hTexRef, unsigned int Flags)
+{
+    TRACE("(%p, %u)\n", hTexRef, Flags);
+    return pcuTexRefSetFlags(hTexRef, Flags);
+}
+
+CUresult WINAPI wine_cuTexRefSetFormat(CUtexref hTexRef, CUarray_format fmt, int NumPackedComponents)
+{
+    TRACE("(%p, %d, %d)\n", hTexRef, fmt, NumPackedComponents);
+    return pcuTexRefSetFormat(hTexRef, fmt, NumPackedComponents);
+}
+
+CUresult WINAPI wine_cuTexRefSetMaxAnisotropy(CUtexref hTexRef, unsigned int maxAniso)
+{
+    TRACE("(%p, %u)\n", hTexRef, maxAniso);
+    return pcuTexRefSetMaxAnisotropy(hTexRef, maxAniso);
+}
+
+CUresult WINAPI wine_cuTexRefSetMipmapFilterMode(CUtexref hTexRef, CUfilter_mode fm)
+{
+    TRACE("(%p, %u)\n", hTexRef, fm);
+    return pcuTexRefSetMipmapFilterMode(hTexRef, fm);
+}
+
+CUresult WINAPI wine_cuTexRefSetMipmapLevelBias(CUtexref hTexRef, float bias)
+{
+    TRACE("(%p, %f)\n", hTexRef, bias);
+    return pcuTexRefSetMipmapLevelBias(hTexRef, bias);
+}
+
+CUresult WINAPI wine_cuTexRefSetMipmapLevelClamp(CUtexref hTexRef, float minMipmapLevelClamp, float maxMipmapLevelClamp)
+{
+    TRACE("(%p, %f, %f)\n", hTexRef, minMipmapLevelClamp, maxMipmapLevelClamp);
+    return pcuTexRefSetMipmapLevelClamp(hTexRef, minMipmapLevelClamp, maxMipmapLevelClamp);
+}
+
+CUresult WINAPI wine_cuTexRefSetMipmappedArray(CUtexref hTexRef, CUmipmappedArray hMipmappedArray, unsigned int Flags)
+{
+    TRACE("(%p, %p, %u)\n", hTexRef, hMipmappedArray, Flags);
+    return pcuTexRefSetMipmappedArray(hTexRef, hMipmappedArray, Flags);
+}
 
 BOOL WINAPI DllMain(HINSTANCE instance, DWORD reason, LPVOID reserved)
 {
+    TRACE("(%p, %u, %p)\n", instance, reason, reserved);
+
     switch (reason)
     {
-        case DLL_WINE_PREATTACH:
-            return FALSE;    /* prefer native version */
         case DLL_PROCESS_ATTACH:
-            DisableThreadLibraryCalls(instance);
+            if (!load_functions()) return FALSE;
+            break;
+        case DLL_PROCESS_DETACH:
+            if (reserved) break;
+            if (cuda_handle) wine_dlclose(cuda_handle, NULL, 0);
+            break;
+        case DLL_THREAD_ATTACH:
+        case DLL_THREAD_DETACH:
+            cuda_process_tls_callbacks(reason);
             break;
     }
 
diff --git a/dlls/nvcuda/nvcuda.h b/dlls/nvcuda/nvcuda.h
new file mode 100644
index 0000000..aaffe13
--- /dev/null
+++ b/dlls/nvcuda/nvcuda.h
@@ -0,0 +1,29 @@
+/*
+ * Copyright (C) 2014-2015 Michael Müller
+ *
+ * This library is free software; you can redistribute it and/or
+ * modify it under the terms of the GNU Lesser General Public
+ * License as published by the Free Software Foundation; either
+ * version 2.1 of the License, or (at your option) any later version.
+ *
+ * This library is distributed in the hope that it will be useful,
+ * but WITHOUT ANY WARRANTY; without even the implied warranty of
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the GNU
+ * Lesser General Public License for more details.
+ *
+ * You should have received a copy of the GNU Lesser General Public
+ * License along with this library; if not, write to the Free Software
+ * Foundation, Inc., 51 Franklin St, Fifth Floor, Boston, MA 02110-1301, USA
+ */
+
+#ifndef __WINE_NVCUDA_H
+#define __WINE_NVCUDA_H
+
+#include "windef.h"
+#include "winbase.h"
+#include "cuda.h"
+
+void cuda_process_tls_callbacks(DWORD reason) DECLSPEC_HIDDEN;
+CUresult cuda_get_table(const void **table, const CUuuid *id, const void *orig_table, CUresult orig_result) DECLSPEC_HIDDEN;
+
+#endif
diff --git a/dlls/nvcuda/nvcuda.spec b/dlls/nvcuda/nvcuda.spec
index 279f7ab..621b8d5 100644
--- a/dlls/nvcuda/nvcuda.spec
+++ b/dlls/nvcuda/nvcuda.spec
@@ -1,36 +1,36 @@
-@ stub cuArray3DCreate
-@ stub cuArray3DCreate_v2
-@ stub cuArray3DGetDescriptor
-@ stub cuArray3DGetDescriptor_v2
-@ stub cuArrayCreate
-@ stub cuArrayCreate_v2
-@ stub cuArrayDestroy
-@ stub cuArrayGetDescriptor
-@ stub cuArrayGetDescriptor_v2
-@ stub cuCtxAttach
-@ stub cuCtxCreate
-@ stub cuCtxCreate_v2
-@ stub cuCtxDestroy
-@ stub cuCtxDestroy_v2
-@ stub cuCtxDetach
-@ stub cuCtxDisablePeerAccess
-@ stub cuCtxEnablePeerAccess
-@ stub cuCtxGetApiVersion
-@ stub cuCtxGetCacheConfig
-@ stub cuCtxGetCurrent
-@ stub cuCtxGetDevice
-@ stub cuCtxGetLimit
-@ stub cuCtxGetSharedMemConfig
-@ stub cuCtxGetStreamPriorityRange
-@ stub cuCtxPopCurrent
-@ stub cuCtxPopCurrent_v2
-@ stub cuCtxPushCurrent
-@ stub cuCtxPushCurrent_v2
-@ stub cuCtxSetCacheConfig
-@ stub cuCtxSetCurrent
-@ stub cuCtxSetLimit
-@ stub cuCtxSetSharedMemConfig
-@ stub cuCtxSynchronize
+@ stdcall cuArray3DCreate(ptr ptr) wine_cuArray3DCreate
+@ stdcall cuArray3DCreate_v2(ptr ptr) wine_cuArray3DCreate_v2
+@ stdcall cuArray3DGetDescriptor(ptr ptr) wine_cuArray3DGetDescriptor
+@ stdcall cuArray3DGetDescriptor_v2(ptr ptr) wine_cuArray3DGetDescriptor_v2
+@ stdcall cuArrayCreate(ptr ptr) wine_cuArrayCreate
+@ stdcall cuArrayCreate_v2(ptr ptr) wine_cuArrayCreate_v2
+@ stdcall cuArrayDestroy(ptr) wine_cuArrayDestroy
+@ stdcall cuArrayGetDescriptor(ptr ptr) wine_cuArrayGetDescriptor
+@ stdcall cuArrayGetDescriptor_v2(ptr ptr) wine_cuArrayGetDescriptor_v2
+@ stdcall cuCtxAttach(ptr long) wine_cuCtxAttach
+@ stdcall cuCtxCreate(ptr long long) wine_cuCtxCreate
+@ stdcall cuCtxCreate_v2(ptr long long) wine_cuCtxCreate_v2
+@ stdcall cuCtxDestroy(ptr) wine_cuCtxDestroy
+@ stdcall cuCtxDestroy_v2(ptr) wine_cuCtxDestroy_v2
+@ stdcall cuCtxDetach(ptr) wine_cuCtxDetach
+@ stdcall cuCtxDisablePeerAccess(ptr) wine_cuCtxDisablePeerAccess
+@ stdcall cuCtxEnablePeerAccess(ptr long) wine_cuCtxEnablePeerAccess
+@ stdcall cuCtxGetApiVersion(ptr ptr) wine_cuCtxGetApiVersion
+@ stdcall cuCtxGetCacheConfig(ptr) wine_cuCtxGetCacheConfig
+@ stdcall cuCtxGetCurrent(ptr) wine_cuCtxGetCurrent
+@ stdcall cuCtxGetDevice(ptr) wine_cuCtxGetDevice
+@ stdcall cuCtxGetLimit(ptr long) wine_cuCtxGetLimit
+@ stdcall cuCtxGetSharedMemConfig(ptr) wine_cuCtxGetSharedMemConfig
+@ stdcall cuCtxGetStreamPriorityRange(ptr ptr) wine_cuCtxGetStreamPriorityRange
+@ stdcall cuCtxPopCurrent(ptr) wine_cuCtxPopCurrent
+@ stdcall cuCtxPopCurrent_v2(ptr) wine_cuCtxPopCurrent_v2
+@ stdcall cuCtxPushCurrent(ptr) wine_cuCtxPushCurrent
+@ stdcall cuCtxPushCurrent_v2(ptr) wine_cuCtxPushCurrent_v2
+@ stdcall cuCtxSetCacheConfig(long) wine_cuCtxSetCacheConfig
+@ stdcall cuCtxSetCurrent(ptr) wine_cuCtxSetCurrent
+@ stdcall cuCtxSetLimit(long long) wine_cuCtxSetLimit
+@ stdcall cuCtxSetSharedMemConfig(long) wine_cuCtxSetSharedMemConfig
+@ stdcall cuCtxSynchronize() wine_cuCtxSynchronize
 @ stub cuD3D10CtxCreate
 @ stub cuD3D10CtxCreateOnDevice
 @ stub cuD3D10CtxCreate_v2
@@ -84,225 +84,225 @@
 @ stub cuD3D9UnmapVertexBuffer
 @ stub cuD3D9UnregisterResource
 @ stub cuD3D9UnregisterVertexBuffer
-@ stub cuDeviceCanAccessPeer
-@ stub cuDeviceComputeCapability
-@ stub cuDeviceGet
-@ stub cuDeviceGetAttribute
-@ stub cuDeviceGetByPCIBusId
-@ stub cuDeviceGetCount
-@ stub cuDeviceGetName
-@ stub cuDeviceGetPCIBusId
-@ stub cuDeviceGetProperties
-@ stub cuDeviceTotalMem
-@ stub cuDeviceTotalMem_v2
-@ stub cuDriverGetVersion
-@ stub cuEventCreate
-@ stub cuEventDestroy
-@ stub cuEventDestroy_v2
-@ stub cuEventElapsedTime
-@ stub cuEventQuery
-@ stub cuEventRecord
-@ stub cuEventSynchronize
-@ stub cuFuncGetAttribute
-@ stub cuFuncSetBlockShape
-@ stub cuFuncSetCacheConfig
-@ stub cuFuncSetSharedMemConfig
-@ stub cuFuncSetSharedSize
-@ stub cuGLCtxCreate
-@ stub cuGLCtxCreate_v2
-@ stub cuGLGetDevices
-@ stub cuGLInit
-@ stub cuGLMapBufferObject
-@ stub cuGLMapBufferObjectAsync
-@ stub cuGLMapBufferObjectAsync_v2
-@ stub cuGLMapBufferObject_v2
-@ stub cuGLRegisterBufferObject
-@ stub cuGLSetBufferObjectMapFlags
-@ stub cuGLUnmapBufferObject
-@ stub cuGLUnmapBufferObjectAsync
-@ stub cuGLUnregisterBufferObject
-@ stub cuGetErrorName
-@ stub cuGetErrorString
-@ stub cuGetExportTable
+@ stdcall cuDeviceCanAccessPeer(ptr long long) wine_cuDeviceCanAccessPeer
+@ stdcall cuDeviceComputeCapability(ptr ptr long) wine_cuDeviceComputeCapability
+@ stdcall cuDeviceGet(ptr long) wine_cuDeviceGet
+@ stdcall cuDeviceGetAttribute(ptr long long) wine_cuDeviceGetAttribute
+@ stdcall cuDeviceGetByPCIBusId(ptr str) wine_cuDeviceGetByPCIBusId
+@ stdcall cuDeviceGetCount(ptr) wine_cuDeviceGetCount
+@ stdcall cuDeviceGetName(ptr long long) wine_cuDeviceGetName
+@ stdcall cuDeviceGetPCIBusId(ptr long long) wine_cuDeviceGetPCIBusId
+@ stdcall cuDeviceGetProperties(ptr long) wine_cuDeviceGetProperties
+@ stdcall cuDeviceTotalMem(ptr long) wine_cuDeviceTotalMem
+@ stdcall cuDeviceTotalMem_v2(ptr long) wine_cuDeviceTotalMem_v2
+@ stdcall cuDriverGetVersion(ptr) wine_cuDriverGetVersion
+@ stdcall cuEventCreate(ptr long) wine_cuEventCreate
+@ stdcall cuEventDestroy(ptr) wine_cuEventDestroy
+@ stdcall cuEventDestroy_v2(ptr) wine_cuEventDestroy_v2
+@ stdcall cuEventElapsedTime(ptr ptr ptr) wine_cuEventElapsedTime
+@ stdcall cuEventQuery(ptr) wine_cuEventQuery
+@ stdcall cuEventRecord(ptr ptr) wine_cuEventRecord
+@ stdcall cuEventSynchronize(ptr) wine_cuEventSynchronize
+@ stdcall cuFuncGetAttribute(ptr long ptr) wine_cuFuncGetAttribute
+@ stdcall cuFuncSetBlockShape(ptr long long long) wine_cuFuncSetBlockShape
+@ stdcall cuFuncSetCacheConfig(ptr long) wine_cuFuncSetCacheConfig
+@ stdcall cuFuncSetSharedMemConfig(ptr long) wine_cuFuncSetSharedMemConfig
+@ stdcall cuFuncSetSharedSize(ptr long) wine_cuFuncSetSharedSize
+@ stdcall cuGLCtxCreate(ptr long long) wine_cuGLCtxCreate
+@ stdcall cuGLCtxCreate_v2(ptr long long) wine_cuGLCtxCreate_v2
+@ stdcall cuGLGetDevices(ptr ptr long long) wine_cuGLGetDevices
+@ stdcall cuGLInit() wine_cuGLInit
+@ stdcall cuGLMapBufferObject(ptr ptr long) wine_cuGLMapBufferObject
+@ stdcall cuGLMapBufferObjectAsync(ptr ptr long ptr) wine_cuGLMapBufferObjectAsync
+@ stdcall cuGLMapBufferObjectAsync_v2(ptr ptr long ptr) wine_cuGLMapBufferObjectAsync_v2
+@ stdcall cuGLMapBufferObject_v2(ptr ptr long) wine_cuGLMapBufferObject_v2
+@ stdcall cuGLRegisterBufferObject(long) wine_cuGLRegisterBufferObject
+@ stdcall cuGLSetBufferObjectMapFlags(long long) wine_cuGLSetBufferObjectMapFlags
+@ stdcall cuGLUnmapBufferObject(long) wine_cuGLUnmapBufferObject
+@ stdcall cuGLUnmapBufferObjectAsync(long ptr) wine_cuGLUnmapBufferObjectAsync
+@ stdcall cuGLUnregisterBufferObject(long) wine_cuGLUnregisterBufferObject
+@ stdcall cuGetErrorName(long ptr) wine_cuGetErrorName
+@ stdcall cuGetErrorString(long ptr) wine_cuGetErrorString
+@ stdcall cuGetExportTable(ptr ptr) wine_cuGetExportTable
 @ stub cuGraphicsD3D10RegisterResource
 @ stub cuGraphicsD3D11RegisterResource
 @ stub cuGraphicsD3D9RegisterResource
-@ stub cuGraphicsGLRegisterBuffer
-@ stub cuGraphicsGLRegisterImage
-@ stub cuGraphicsMapResources
-@ stub cuGraphicsResourceGetMappedMipmappedArray
-@ stub cuGraphicsResourceGetMappedPointer
-@ stub cuGraphicsResourceGetMappedPointer_v2
-@ stub cuGraphicsResourceSetMapFlags
-@ stub cuGraphicsSubResourceGetMappedArray
-@ stub cuGraphicsUnmapResources
-@ stub cuGraphicsUnregisterResource
-@ stub cuInit
-@ stub cuIpcCloseMemHandle
-@ stub cuIpcGetEventHandle
-@ stub cuIpcGetMemHandle
-@ stub cuIpcOpenEventHandle
-@ stub cuIpcOpenMemHandle
-@ stub cuLaunch
-@ stub cuLaunchGrid
-@ stub cuLaunchGridAsync
-@ stub cuLaunchKernel
-@ stub cuLinkAddData
+@ stdcall cuGraphicsGLRegisterBuffer(ptr long long) wine_cuGraphicsGLRegisterBuffer
+@ stdcall cuGraphicsGLRegisterImage(ptr long long long) wine_cuGraphicsGLRegisterImage
+@ stdcall cuGraphicsMapResources(long ptr ptr) wine_cuGraphicsMapResources
+@ stdcall cuGraphicsResourceGetMappedMipmappedArray(ptr ptr) wine_cuGraphicsResourceGetMappedMipmappedArray
+@ stdcall cuGraphicsResourceGetMappedPointer(ptr ptr ptr) wine_cuGraphicsResourceGetMappedPointer
+@ stdcall cuGraphicsResourceGetMappedPointer_v2(ptr ptr ptr) wine_cuGraphicsResourceGetMappedPointer_v2
+@ stdcall cuGraphicsResourceSetMapFlags(ptr long) wine_cuGraphicsResourceSetMapFlags
+@ stdcall cuGraphicsSubResourceGetMappedArray(ptr ptr long long) wine_cuGraphicsSubResourceGetMappedArray
+@ stdcall cuGraphicsUnmapResources(long ptr ptr) wine_cuGraphicsUnmapResources
+@ stdcall cuGraphicsUnregisterResource(ptr) wine_cuGraphicsUnregisterResource
+@ stdcall cuInit(long) wine_cuInit
+@ stdcall cuIpcCloseMemHandle(long) wine_cuIpcCloseMemHandle
+@ stdcall cuIpcGetEventHandle(ptr ptr) wine_cuIpcGetEventHandle
+@ stdcall cuIpcGetMemHandle(ptr long) wine_cuIpcGetMemHandle
+@ stdcall cuIpcOpenEventHandle(ptr ptr) wine_cuIpcOpenEventHandle
+@ stdcall cuIpcOpenMemHandle(ptr ptr long) wine_cuIpcOpenMemHandle
+@ stdcall cuLaunch(ptr) wine_cuLaunch
+@ stdcall cuLaunchGrid(ptr long long) wine_cuLaunchGrid
+@ stdcall cuLaunchGridAsync(ptr long long ptr) wine_cuLaunchGridAsync
+@ stdcall cuLaunchKernel(ptr long long long long long long long ptr ptr ptr) wine_cuLaunchKernel
+@ stdcall cuLinkAddData(ptr long ptr long str long ptr ptr) wine_cuLinkAddData
 @ stub cuLinkAddFile
-@ stub cuLinkComplete
-@ stub cuLinkCreate
-@ stub cuLinkDestroy
-@ stub cuMemAlloc
-@ stub cuMemAllocHost
-@ stub cuMemAllocHost_v2
-@ stub cuMemAllocManaged
-@ stub cuMemAllocPitch
-@ stub cuMemAllocPitch_v2
-@ stub cuMemAlloc_v2
-@ stub cuMemFree
-@ stub cuMemFreeHost
-@ stub cuMemFree_v2
-@ stub cuMemGetAddressRange
-@ stub cuMemGetAddressRange_v2
-@ stub cuMemGetInfo
-@ stub cuMemGetInfo_v2
-@ stub cuMemHostAlloc
-@ stub cuMemHostGetDevicePointer
-@ stub cuMemHostGetDevicePointer_v2
-@ stub cuMemHostGetFlags
-@ stub cuMemHostRegister
-@ stub cuMemHostUnregister
-@ stub cuMemcpy
-@ stub cuMemcpy2D
-@ stub cuMemcpy2DAsync
-@ stub cuMemcpy2DAsync_v2
-@ stub cuMemcpy2DUnaligned
-@ stub cuMemcpy2DUnaligned_v2
-@ stub cuMemcpy2D_v2
-@ stub cuMemcpy3D
-@ stub cuMemcpy3DAsync
-@ stub cuMemcpy3DAsync_v2
-@ stub cuMemcpy3DPeer
-@ stub cuMemcpy3DPeerAsync
-@ stub cuMemcpy3D_v2
-@ stub cuMemcpyAsync
-@ stub cuMemcpyAtoA
-@ stub cuMemcpyAtoA_v2
-@ stub cuMemcpyAtoD
-@ stub cuMemcpyAtoD_v2
-@ stub cuMemcpyAtoH
-@ stub cuMemcpyAtoHAsync
-@ stub cuMemcpyAtoHAsync_v2
-@ stub cuMemcpyAtoH_v2
-@ stub cuMemcpyDtoA
-@ stub cuMemcpyDtoA_v2
-@ stub cuMemcpyDtoD
-@ stub cuMemcpyDtoDAsync
-@ stub cuMemcpyDtoDAsync_v2
-@ stub cuMemcpyDtoD_v2
-@ stub cuMemcpyDtoH
-@ stub cuMemcpyDtoHAsync
-@ stub cuMemcpyDtoHAsync_v2
-@ stub cuMemcpyDtoH_v2
-@ stub cuMemcpyHtoA
-@ stub cuMemcpyHtoAAsync
-@ stub cuMemcpyHtoAAsync_v2
-@ stub cuMemcpyHtoA_v2
-@ stub cuMemcpyHtoD
-@ stub cuMemcpyHtoDAsync
-@ stub cuMemcpyHtoDAsync_v2
-@ stub cuMemcpyHtoD_v2
-@ stub cuMemcpyPeer
-@ stub cuMemcpyPeerAsync
-@ stub cuMemsetD16
-@ stub cuMemsetD16Async
-@ stub cuMemsetD16_v2
-@ stub cuMemsetD2D16
-@ stub cuMemsetD2D16Async
-@ stub cuMemsetD2D16_v2
-@ stub cuMemsetD2D32
-@ stub cuMemsetD2D32Async
-@ stub cuMemsetD2D32_v2
-@ stub cuMemsetD2D8
-@ stub cuMemsetD2D8Async
-@ stub cuMemsetD2D8_v2
-@ stub cuMemsetD32
-@ stub cuMemsetD32Async
-@ stub cuMemsetD32_v2
-@ stub cuMemsetD8
-@ stub cuMemsetD8Async
-@ stub cuMemsetD8_v2
-@ stub cuMipmappedArrayCreate
-@ stub cuMipmappedArrayDestroy
-@ stub cuMipmappedArrayGetLevel
-@ stub cuModuleGetFunction
-@ stub cuModuleGetGlobal
-@ stub cuModuleGetGlobal_v2
-@ stub cuModuleGetSurfRef
-@ stub cuModuleGetTexRef
+@ stdcall cuLinkComplete(ptr ptr ptr) wine_cuLinkComplete
+@ stdcall cuLinkCreate(long ptr ptr ptr) wine_cuLinkCreate
+@ stdcall cuLinkDestroy(ptr) wine_cuLinkDestroy
+@ stdcall cuMemAlloc(ptr long) wine_cuMemAlloc
+@ stdcall cuMemAllocHost(ptr long) wine_cuMemAllocHost
+@ stdcall cuMemAllocHost_v2(ptr long) wine_cuMemAllocHost_v2
+@ stdcall cuMemAllocManaged(ptr long long) wine_cuMemAllocManaged
+@ stdcall cuMemAllocPitch(ptr ptr long long long) wine_cuMemAllocPitch
+@ stdcall cuMemAllocPitch_v2(ptr ptr long long long) wine_cuMemAllocPitch_v2
+@ stdcall cuMemAlloc_v2(ptr long) wine_cuMemAlloc_v2
+@ stdcall cuMemFree(long) wine_cuMemFree
+@ stdcall cuMemFreeHost(ptr) wine_cuMemFreeHost
+@ stdcall cuMemFree_v2(long) wine_cuMemFree_v2
+@ stdcall cuMemGetAddressRange(ptr ptr long) wine_cuMemGetAddressRange
+@ stdcall cuMemGetAddressRange_v2(ptr ptr long) wine_cuMemGetAddressRange_v2
+@ stdcall cuMemGetInfo(ptr ptr) wine_cuMemGetInfo
+@ stdcall cuMemGetInfo_v2(ptr ptr) wine_cuMemGetInfo_v2
+@ stdcall cuMemHostAlloc(ptr long long) wine_cuMemHostAlloc
+@ stdcall cuMemHostGetDevicePointer(ptr ptr long) wine_cuMemHostGetDevicePointer
+@ stdcall cuMemHostGetDevicePointer_v2(ptr ptr long) wine_cuMemHostGetDevicePointer_v2
+@ stdcall cuMemHostGetFlags(ptr ptr) wine_cuMemHostGetFlags
+@ stdcall cuMemHostRegister(ptr long long) wine_cuMemHostRegister
+@ stdcall cuMemHostUnregister(ptr) wine_cuMemHostUnregister
+@ stdcall cuMemcpy2D(ptr) wine_cuMemcpy2D
+@ stdcall cuMemcpy2DAsync(ptr ptr) wine_cuMemcpy2DAsync
+@ stdcall cuMemcpy2DAsync_v2(ptr ptr) wine_cuMemcpy2DAsync_v2
+@ stdcall cuMemcpy2DUnaligned(ptr) wine_cuMemcpy2DUnaligned
+@ stdcall cuMemcpy2DUnaligned_v2(ptr) wine_cuMemcpy2DUnaligned_v2
+@ stdcall cuMemcpy2D_v2(ptr) wine_cuMemcpy2D_v2
+@ stdcall cuMemcpy3D(ptr) wine_cuMemcpy3D
+@ stdcall cuMemcpy3DAsync(ptr ptr) wine_cuMemcpy3DAsync
+@ stdcall cuMemcpy3DAsync_v2(ptr ptr) wine_cuMemcpy3DAsync_v2
+@ stdcall cuMemcpy3DPeer(ptr) wine_cuMemcpy3DPeer
+@ stdcall cuMemcpy3DPeerAsync(ptr ptr) wine_cuMemcpy3DPeerAsync
+@ stdcall cuMemcpy3D_v2(ptr) wine_cuMemcpy3D_v2
+@ stdcall cuMemcpy(long long long) wine_cuMemcpy
+@ stdcall cuMemcpyAsync(long long long ptr) wine_cuMemcpyAsync
+@ stdcall cuMemcpyAtoA(ptr long ptr long long) wine_cuMemcpyAtoA
+@ stdcall cuMemcpyAtoA_v2(ptr long ptr long long) wine_cuMemcpyAtoA_v2
+@ stdcall cuMemcpyAtoD(long ptr long long) wine_cuMemcpyAtoD
+@ stdcall cuMemcpyAtoD_v2(long ptr long long) wine_cuMemcpyAtoD_v2
+@ stdcall cuMemcpyAtoH(ptr ptr long long) wine_cuMemcpyAtoH
+@ stdcall cuMemcpyAtoHAsync(ptr ptr long long ptr) wine_cuMemcpyAtoHAsync
+@ stdcall cuMemcpyAtoHAsync_v2(ptr ptr long long ptr) wine_cuMemcpyAtoHAsync_v2
+@ stdcall cuMemcpyAtoH_v2(ptr ptr long long) wine_cuMemcpyAtoH_v2
+@ stdcall cuMemcpyDtoA(ptr long long long) wine_cuMemcpyDtoA
+@ stdcall cuMemcpyDtoA_v2(ptr long long long) wine_cuMemcpyDtoA_v2
+@ stdcall cuMemcpyDtoD(long long long) wine_cuMemcpyDtoD
+@ stdcall cuMemcpyDtoDAsync(long long long ptr) wine_cuMemcpyDtoDAsync
+@ stdcall cuMemcpyDtoDAsync_v2(long long long ptr) wine_cuMemcpyDtoDAsync_v2
+@ stdcall cuMemcpyDtoD_v2(long long long) wine_cuMemcpyDtoD_v2
+@ stdcall cuMemcpyDtoH(ptr long long) wine_cuMemcpyDtoH
+@ stdcall cuMemcpyDtoHAsync(ptr long long ptr) wine_cuMemcpyDtoHAsync
+@ stdcall cuMemcpyDtoHAsync_v2(ptr long long ptr) wine_cuMemcpyDtoHAsync_v2
+@ stdcall cuMemcpyDtoH_v2(ptr long long) wine_cuMemcpyDtoH_v2
+@ stdcall cuMemcpyHtoA(ptr long ptr long) wine_cuMemcpyHtoA
+@ stdcall cuMemcpyHtoAAsync(ptr long ptr long ptr) wine_cuMemcpyHtoAAsync
+@ stdcall cuMemcpyHtoAAsync_v2(ptr long ptr long ptr) wine_cuMemcpyHtoAAsync_v2
+@ stdcall cuMemcpyHtoA_v2(ptr long ptr long) wine_cuMemcpyHtoA_v2
+@ stdcall cuMemcpyHtoD(long ptr long) wine_cuMemcpyHtoD
+@ stdcall cuMemcpyHtoDAsync(long ptr long ptr) wine_cuMemcpyHtoDAsync
+@ stdcall cuMemcpyHtoDAsync_v2(long ptr long ptr) wine_cuMemcpyHtoDAsync_v2
+@ stdcall cuMemcpyHtoD_v2(long ptr long) wine_cuMemcpyHtoD_v2
+@ stdcall cuMemcpyPeer(long ptr long ptr long) wine_cuMemcpyPeer
+@ stdcall cuMemcpyPeerAsync(long ptr long ptr long ptr) wine_cuMemcpyPeerAsync
+@ stdcall cuMemsetD16(long long long) wine_cuMemsetD16
+@ stdcall cuMemsetD16Async(long long long ptr) wine_cuMemsetD16Async
+@ stdcall cuMemsetD16_v2(long long long) wine_cuMemsetD16_v2
+@ stdcall cuMemsetD2D16(long long long long long) wine_cuMemsetD2D16
+@ stdcall cuMemsetD2D16Async(long long long long long ptr) wine_cuMemsetD2D16Async
+@ stdcall cuMemsetD2D16_v2(long long long long long) wine_cuMemsetD2D16_v2
+@ stdcall cuMemsetD2D32(long long long long long) wine_cuMemsetD2D32
+@ stdcall cuMemsetD2D32Async(long long long long long ptr) wine_cuMemsetD2D32Async
+@ stdcall cuMemsetD2D32_v2(long long long long long) wine_cuMemsetD2D32_v2
+@ stdcall cuMemsetD2D8(long long long long long) wine_cuMemsetD2D8
+@ stdcall cuMemsetD2D8Async(long long long long long ptr) wine_cuMemsetD2D8Async
+@ stdcall cuMemsetD2D8_v2(long long long long long) wine_cuMemsetD2D8_v2
+@ stdcall cuMemsetD32(long long long) wine_cuMemsetD32
+@ stdcall cuMemsetD32Async(long long long ptr) wine_cuMemsetD32Async
+@ stdcall cuMemsetD32_v2(long long long) wine_cuMemsetD32_v2
+@ stdcall cuMemsetD8(long long long) wine_cuMemsetD8
+@ stdcall cuMemsetD8Async(long long long ptr) wine_cuMemsetD8Async
+@ stdcall cuMemsetD8_v2(long long long) wine_cuMemsetD8_v2
+@ stdcall cuMipmappedArrayCreate(ptr ptr long) wine_cuMipmappedArrayCreate
+@ stdcall cuMipmappedArrayDestroy(ptr) wine_cuMipmappedArrayDestroy
+@ stdcall cuMipmappedArrayGetLevel(ptr ptr long) wine_cuMipmappedArrayGetLevel
+@ stdcall cuModuleGetFunction(ptr ptr str) wine_cuModuleGetFunction
+@ stdcall cuModuleGetGlobal(ptr ptr ptr str) wine_cuModuleGetGlobal
+@ stdcall cuModuleGetGlobal_v2(ptr ptr ptr str) wine_cuModuleGetGlobal_v2
+@ stdcall cuModuleGetSurfRef(ptr ptr str) wine_cuModuleGetSurfRef
+@ stdcall cuModuleGetTexRef(ptr ptr str) wine_cuModuleGetTexRef
 @ stub cuModuleLoad
-@ stub cuModuleLoadData
-@ stub cuModuleLoadDataEx
-@ stub cuModuleLoadFatBinary
-@ stub cuModuleUnload
-@ stub cuParamSetSize
-@ stub cuParamSetTexRef
-@ stub cuParamSetf
-@ stub cuParamSeti
-@ stub cuParamSetv
-@ stub cuPointerGetAttribute
-@ stub cuPointerSetAttribute
+@ stdcall cuModuleLoadData(ptr ptr) wine_cuModuleLoadData
+@ stdcall cuModuleLoadDataEx(ptr ptr long ptr ptr) wine_cuModuleLoadDataEx
+@ stdcall cuModuleLoadFatBinary(ptr ptr) wine_cuModuleLoadFatBinary
+@ stdcall cuModuleUnload(ptr) wine_cuModuleUnload
+@ stdcall cuParamSetSize(ptr long) wine_cuParamSetSize
+@ stdcall cuParamSetTexRef(ptr long ptr) wine_cuParamSetTexRef
+@ stdcall cuParamSetf(ptr long float) wine_cuParamSetf
+@ stdcall cuParamSeti(ptr long long) wine_cuParamSeti
+@ stdcall cuParamSetv(ptr long ptr long) wine_cuParamSetv
+@ stdcall cuPointerGetAttribute(ptr long long) wine_cuPointerGetAttribute
+@ stdcall cuPointerSetAttribute(ptr long long) wine_cuPointerSetAttribute
 @ stub cuProfilerInitialize
 @ stub cuProfilerStart
 @ stub cuProfilerStop
-@ stub cuStreamAddCallback
-@ stub cuStreamAttachMemAsync
-@ stub cuStreamCreate
-@ stub cuStreamCreateWithPriority
-@ stub cuStreamDestroy
-@ stub cuStreamDestroy_v2
-@ stub cuStreamGetFlags
-@ stub cuStreamGetPriority
-@ stub cuStreamQuery
-@ stub cuStreamSynchronize
-@ stub cuStreamWaitEvent
-@ stub cuSurfObjectCreate
-@ stub cuSurfObjectDestroy
-@ stub cuSurfObjectGetResourceDesc
-@ stub cuSurfRefGetArray
-@ stub cuSurfRefSetArray
-@ stub cuTexObjectCreate
-@ stub cuTexObjectDestroy
-@ stub cuTexObjectGetResourceDesc
-@ stub cuTexObjectGetResourceViewDesc
-@ stub cuTexObjectGetTextureDesc
-@ stub cuTexRefCreate
-@ stub cuTexRefDestroy
-@ stub cuTexRefGetAddress
-@ stub cuTexRefGetAddressMode
-@ stub cuTexRefGetAddress_v2
-@ stub cuTexRefGetArray
-@ stub cuTexRefGetFilterMode
-@ stub cuTexRefGetFlags
-@ stub cuTexRefGetFormat
-@ stub cuTexRefGetMaxAnisotropy
-@ stub cuTexRefGetMipmapFilterMode
-@ stub cuTexRefGetMipmapLevelBias
-@ stub cuTexRefGetMipmapLevelClamp
-@ stub cuTexRefGetMipmappedArray
-@ stub cuTexRefSetAddress
-@ stub cuTexRefSetAddress2D
-@ stub cuTexRefSetAddress2D_v2
-@ stub cuTexRefSetAddress2D_v3
-@ stub cuTexRefSetAddressMode
-@ stub cuTexRefSetAddress_v2
-@ stub cuTexRefSetArray
-@ stub cuTexRefSetFilterMode
-@ stub cuTexRefSetFlags
-@ stub cuTexRefSetFormat
-@ stub cuTexRefSetMaxAnisotropy
-@ stub cuTexRefSetMipmapFilterMode
-@ stub cuTexRefSetMipmapLevelBias
-@ stub cuTexRefSetMipmapLevelClamp
-@ stub cuTexRefSetMipmappedArray
+@ stdcall cuStreamAddCallback(ptr ptr ptr long) wine_cuStreamAddCallback
+@ stdcall cuStreamAttachMemAsync(ptr long long long) wine_cuStreamAttachMemAsync
+@ stdcall cuStreamCreate(ptr long) wine_cuStreamCreate
+@ stdcall cuStreamCreateWithPriority(ptr long long) wine_cuStreamCreateWithPriority
+@ stdcall cuStreamDestroy(ptr) wine_cuStreamDestroy
+@ stdcall cuStreamDestroy_v2(ptr) wine_cuStreamDestroy
+@ stdcall cuStreamGetFlags(ptr ptr) wine_cuStreamGetFlags
+@ stdcall cuStreamGetPriority(ptr ptr) wine_cuStreamGetPriority
+@ stdcall cuStreamQuery(ptr) wine_cuStreamQuery
+@ stdcall cuStreamSynchronize(ptr) wine_cuStreamSynchronize
+@ stdcall cuStreamWaitEvent(ptr ptr long) wine_cuStreamWaitEvent
+@ stdcall cuSurfObjectCreate(ptr ptr) wine_cuSurfObjectCreate
+@ stdcall cuSurfObjectDestroy(int64) wine_cuSurfObjectDestroy
+@ stdcall cuSurfObjectGetResourceDesc(ptr int64) wine_cuSurfObjectGetResourceDesc
+@ stdcall cuSurfRefGetArray(ptr ptr) wine_cuSurfRefGetArray
+@ stdcall cuSurfRefSetArray(ptr ptr long) wine_cuSurfRefSetArray
+@ stdcall cuTexObjectCreate(ptr ptr ptr ptr) wine_cuTexObjectCreate
+@ stdcall cuTexObjectDestroy(int64) wine_cuTexObjectDestroy
+@ stdcall cuTexObjectGetResourceDesc(ptr int64) wine_cuTexObjectGetResourceDesc
+@ stdcall cuTexObjectGetResourceViewDesc(ptr int64) wine_cuTexObjectGetResourceViewDesc
+@ stdcall cuTexObjectGetTextureDesc(ptr int64) wine_cuTexObjectGetTextureDesc
+@ stdcall cuTexRefCreate(ptr) wine_cuTexRefCreate
+@ stdcall cuTexRefDestroy(ptr) wine_cuTexRefDestroy
+@ stdcall cuTexRefGetAddress(ptr ptr) wine_cuTexRefGetAddress
+@ stdcall cuTexRefGetAddressMode(ptr ptr long) wine_cuTexRefGetAddressMode
+@ stdcall cuTexRefGetAddress_v2(ptr ptr) wine_cuTexRefGetAddress_v2
+@ stdcall cuTexRefGetArray(ptr ptr) wine_cuTexRefGetArray
+@ stdcall cuTexRefGetFilterMode(ptr ptr) wine_cuTexRefGetFilterMode
+@ stdcall cuTexRefGetFlags(ptr ptr) wine_cuTexRefGetFlags
+@ stdcall cuTexRefGetFormat(ptr ptr ptr) wine_cuTexRefGetFormat
+@ stdcall cuTexRefGetMaxAnisotropy(ptr ptr) wine_cuTexRefGetMaxAnisotropy
+@ stdcall cuTexRefGetMipmapFilterMode(ptr ptr) wine_cuTexRefGetMipmapFilterMode
+@ stdcall cuTexRefGetMipmapLevelBias(ptr ptr) wine_cuTexRefGetMipmapLevelBias
+@ stdcall cuTexRefGetMipmapLevelClamp(ptr ptr ptr) wine_cuTexRefGetMipmapLevelClamp
+@ stdcall cuTexRefGetMipmappedArray(ptr ptr) wine_cuTexRefGetMipmappedArray
+@ stdcall cuTexRefSetAddress2D(ptr ptr long long) wine_cuTexRefSetAddress2D
+@ stdcall cuTexRefSetAddress2D_v2(ptr ptr long long) wine_cuTexRefSetAddress2D_v2
+@ stdcall cuTexRefSetAddress2D_v3(ptr ptr long long) wine_cuTexRefSetAddress2D_v3
+@ stdcall cuTexRefSetAddress(ptr ptr long long) wine_cuTexRefSetAddress
+@ stdcall cuTexRefSetAddressMode(ptr long long) wine_cuTexRefSetAddressMode
+@ stdcall cuTexRefSetAddress_v2(ptr ptr long long) wine_cuTexRefSetAddress_v2
+@ stdcall cuTexRefSetArray(ptr ptr long) wine_cuTexRefSetArray
+@ stdcall cuTexRefSetFilterMode(ptr long) wine_cuTexRefSetFilterMode
+@ stdcall cuTexRefSetFlags(ptr long) wine_cuTexRefSetFlags
+@ stdcall cuTexRefSetFormat(ptr long long) wine_cuTexRefSetFormat
+@ stdcall cuTexRefSetMaxAnisotropy(ptr long) wine_cuTexRefSetMaxAnisotropy
+@ stdcall cuTexRefSetMipmapFilterMode(ptr long) wine_cuTexRefSetMipmapFilterMode
+@ stdcall cuTexRefSetMipmapLevelBias(ptr float) wine_cuTexRefSetMipmapLevelBias
+@ stdcall cuTexRefSetMipmapLevelClamp(ptr float float) wine_cuTexRefSetMipmapLevelClamp
+@ stdcall cuTexRefSetMipmappedArray(ptr ptr long) wine_cuTexRefSetMipmappedArray
 @ stub cuWGLGetDevice
diff --git a/dlls/nvcuda/tests/Makefile.in b/dlls/nvcuda/tests/Makefile.in
new file mode 100644
index 0000000..81f9290
--- /dev/null
+++ b/dlls/nvcuda/tests/Makefile.in
@@ -0,0 +1,4 @@
+TESTDLL   = nvcuda.dll
+
+C_SRCS = \
+	nvcuda.c
diff --git a/dlls/nvcuda/tests/nvcuda.c b/dlls/nvcuda/tests/nvcuda.c
new file mode 100644
index 0000000..fc8f300
--- /dev/null
+++ b/dlls/nvcuda/tests/nvcuda.c
@@ -0,0 +1,168 @@
+/*
+ * Copyright (C) 2014-2015 Michael Müller
+ *
+ * This library is free software; you can redistribute it and/or
+ * modify it under the terms of the GNU Lesser General Public
+ * License as published by the Free Software Foundation; either
+ * version 2.1 of the License, or (at your option) any later version.
+ *
+ * This library is distributed in the hope that it will be useful,
+ * but WITHOUT ANY WARRANTY; without even the implied warranty of
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the GNU
+ * Lesser General Public License for more details.
+ *
+ * You should have received a copy of the GNU Lesser General Public
+ * License along with this library; if not, write to the Free Software
+ * Foundation, Inc., 51 Franklin St, Fifth Floor, Boston, MA 02110-1301, USA
+ */
+
+#include <stdarg.h>
+
+#include "windef.h"
+#include "winbase.h"
+#include "winerror.h"
+#include "cuda.h"
+
+#include "wine/test.h"
+
+static CUresult (WINAPI *pcuInit)(unsigned int);
+static CUresult (WINAPI *pcuGetExportTable)(const void**, const CUuuid*);
+
+static const CUuuid UUID_TlsNotifyInterface = {{0x19, 0x5B, 0xCB, 0xF4, 0xD6, 0x7D, 0x02, 0x4A,
+                                                0xAC, 0xC5, 0x1D, 0x29, 0xCE, 0xA6, 0x31, 0xAE}};
+
+struct TlsNotifyInterface_table
+{
+    int size;
+    CUresult (WINAPI *Set)(void **handle, void *callback, void *data);
+    CUresult (WINAPI *Remove)(void *handle, void *param1);
+};
+
+static BOOL init(void)
+{
+    HMODULE nvcuda = LoadLibraryA("nvcuda.dll");
+    if (!nvcuda)
+    {
+        skip("Could not load nvcuda.dll\n");
+        return FALSE;
+    }
+
+    pcuInit           = (void *)GetProcAddress(nvcuda, "cuInit");
+    pcuGetExportTable = (void *)GetProcAddress(nvcuda, "cuGetExportTable");
+
+    if (!pcuInit)
+    {
+        win_skip("Failed to get entry point for cuInit.\n");
+        return FALSE;
+    }
+
+    if (pcuInit(0))
+    {
+        skip("Failed to initialize cuda.\n");
+        return FALSE;
+    }
+
+    return TRUE;
+}
+
+struct tls_test_data
+{
+    int count;
+    DWORD reason;
+    DWORD threadid;
+};
+
+static void CDECL tls_callback_test(DWORD reason, void *data)
+{
+    struct tls_test_data *test_data = data;
+    trace("reason: %d, data: %p\n", reason, data);
+
+    test_data->count++;
+    test_data->reason = reason;
+    test_data->threadid = GetCurrentThreadId();
+}
+
+static DWORD WINAPI test_thread(LPVOID data)
+{
+    struct tls_test_data *test_data = data;
+    ok(test_data->count == 0, "Expected 0 callback executions, got %d\n", test_data->count);
+
+    /* do nothing */
+    return 0;
+}
+
+static void test_TlsNotifyInterface(void)
+{
+    const struct TlsNotifyInterface_table *iface;
+    struct tls_test_data test_data;
+    DWORD threadid, thread_res;
+    void *handle, *handle2;
+    HANDLE thread;
+    CUresult res;
+
+    if (!pcuGetExportTable)
+    {
+        win_skip("cuGetExportTable export not found.\n");
+        return;
+    }
+
+    if (pcuGetExportTable((const void **)&iface, &UUID_TlsNotifyInterface))
+    {
+        win_skip("TlsNotifyInterface not supported.\n");
+        return;
+    }
+
+    ok(iface->size == sizeof(struct TlsNotifyInterface_table),
+       "Size mismatch for TlsNotifyInterface, expected %u, got %u\n",
+       (unsigned int)sizeof(struct TlsNotifyInterface_table), iface->size);
+
+    /* Test adding and removing a TLS callback */
+    memset(&test_data, 0x00, sizeof(test_data));
+    res = iface->Set(&handle, &tls_callback_test, &test_data);
+    ok(!res, "Failed to set TLS callback, got error %d\n", res);
+    res = iface->Remove(handle, NULL);
+    ok(!res, "Failed to remove TLS callback, got error %d\n", res);
+    ok(test_data.count == 0, "Expected 0 callback execution, got %d\n", test_data.count);
+
+    /* Test TLS callback with a thread */
+    memset(&test_data, 0x00, sizeof(test_data));
+    res = iface->Set(&handle, &tls_callback_test, &test_data);
+    ok(!res, "Failed to set TLS callback, got error %d\n", res);
+    thread = CreateThread(NULL, 0, test_thread, &test_data, 0, &threadid);
+    ok(thread != NULL, "Failed to create Thread, error: %d\n", GetLastError());
+    thread_res = WaitForSingleObject(thread, 2000);
+    ok(thread_res == WAIT_OBJECT_0, "Waiting for thread failed: %d\n", thread_res);
+    ok(test_data.count == 1, "Expected 1 callback execution, got %d\n", test_data.count);
+    ok(test_data.reason == 0, "Expected reason 0, got %d\n", test_data.reason);
+    ok(test_data.threadid == threadid, "Expected thread id %d, got %d\n", threadid, test_data.threadid);
+    res = iface->Remove(handle, NULL);
+    ok(!res, "Failed to remove TLS callback, got error %d\n", res);
+
+    /* Test adding the same callback twice */
+    res = iface->Set(&handle, &tls_callback_test, &test_data);
+    ok(!res, "Failed to set first TLS callback, got error %d\n", res);
+    res = iface->Set(&handle2, &tls_callback_test, &test_data);
+    ok(!res, "Failed to set second TLS callback, got error %d\n", res);
+    ok(handle != handle2, "expected handle and handle2 to be different\n");
+    res = iface->Remove(handle, NULL);
+    ok(!res, "Failed to remove first TLS callback, got error %d\n", res);
+    res = iface->Remove(handle2, NULL);
+    ok(!res, "Failed to remove second TLS callback, got error %d\n", res);
+
+    /* removing an invalid callback handle causes a crash on Windows */
+    if (0)
+    {
+        res = iface->Remove(handle, NULL);
+        ok(0, "Failed to remove TLS callback, got error %d\n", res);
+        res = iface->Remove(NULL, NULL);
+        ok(0, "Failed to remove TLS callback, got error %d\n", res);
+    }
+}
+
+START_TEST( nvcuda )
+{
+    if (!init())
+        return;
+
+    test_TlsNotifyInterface();
+}
-- 
1.9.1

