cuda内核执行以及运行时和驱动程序 API 之间的互操作性

cuda内核执行以及运行时和驱动程序 API 之间的互操作性

17.3. 内核执行

cuLaunchKernel()启动具有给定执行配置的内核。

参数作为指针数组（在 cuLaunchKernel()的最后一个参数旁边）传递，其中第 n 个指针对应于第 n 个参数并指向从中复制参数的内存区域，或者作为额外选项之一（最后一个参数 cuLaunchKernel()）。

当参数作为额外选项（CU_LAUNCH_PARAM_BUFFER_POINTER选项）传递时，它们将作为指向单个缓冲区的指针传递，其中通过匹配设备代码中每个参数类型的对齐要求，假定参数相对于彼此正确偏移。

对于所有其他基本类型，设备代码中的对齐要求与主机代码中的对齐要求相匹配，因此可以使用 __alignof()获得。唯一的例外是当主机编译器对齐double和long long（在 64 位系统long上）在单字边界而不是双字边界上对齐（例如，使用gcc 的编译-mno-align-double标志），因为在设备代码中，这些类型始终在两个字边界上对齐。

CUdeviceptr是一个整数，但表示一个指针，因此其对齐要求为__alignof(void*) 。

下面的代码示例使用宏 （ALIGN_UP()） 调整每个参数的偏移量以满足其对齐要求，并使用另一个宏 （ADD_TO_PARAM_BUFFER()） 将每个参数添加到传递给CU_LAUNCH_PARAM_BUFFER_POINTER选项的参数缓冲区。

#define ALIGN_UP(offset, alignment) \
      (offset) = ((offset) + (alignment) - 1) & ~((alignment) - 1)
char paramBuffer[1024];
size_t paramBufferSize = 0;
#define ADD_TO_PARAM_BUFFER(value, alignment)                   \
    do {                                                        \
        paramBufferSize = ALIGN_UP(paramBufferSize, alignment); \
        memcpy(paramBuffer + paramBufferSize,                   \
               &(value), sizeof(value));                        \
        paramBufferSize += sizeof(value);                       \
    } while (0)
int i;
ADD_TO_PARAM_BUFFER(i, __alignof(i));
float4 f4;
ADD_TO_PARAM_BUFFER(f4, 16); // float4's alignment is 16
char c;
ADD_TO_PARAM_BUFFER(c, __alignof(c));
float f;
ADD_TO_PARAM_BUFFER(f, __alignof(f));
CUdeviceptr devPtr;
ADD_TO_PARAM_BUFFER(devPtr, __alignof(devPtr));
float2 f2;
ADD_TO_PARAM_BUFFER(f2, 8); // float2's alignment is 8
void* extra[] = {
    CU_LAUNCH_PARAM_BUFFER_POINTER, paramBuffer,
    CU_LAUNCH_PARAM_BUFFER_SIZE,    &paramBufferSize,
    CU_LAUNCH_PARAM_END
cuLaunchKernel(cuFunction,
               blockWidth, blockHeight, blockDepth,
               gridWidth, gridHeight, gridDepth,
               0, 0, 0, extra);

结构的对齐要求等于其字段的对齐要求中的最大值。因此，包含内置向量类型的结构的对齐要求CUdeviceptr或非对齐double和 long long的对齐要求在设备代码和主机代码之间可能有所不同。这种结构也可以以不同的方式填充。例如，以下结构在主机代码中根本不填充，但在设备代码中填充，f字段后有 12 个字节，因为字段的对齐要求f4为 16。

typedef struct {
    float  f;
    float4 f4;
} myStruct;

17.4. 运行时和驱动程序 API 之间的互操作性

应用程序可以将运行时 API 代码与驱动程序 API 代码混合使用。

如果通过驱动程序 API 创建上下文并将其设置为当前上下文，则后续运行时调用将选取此上下文，而不是创建新上下文。

如果运行时已初始化（隐式如 CUDA 运行时 中所述），则cuCtxGetCurrent()可用于检索初始化期间创建的上下文。后续驱动程序 API 调用可以使用此上下文。

从运行时隐式创建的上下文称为 主上下文 （请参阅 初始化 ）。可以使用 主要上下文管理 功能从驱动程序 API 对其进行管理。

可以使用任一 API 分配和释放设备内存。CUdeviceptr可以转换为常规指针，反之亦然：

CUdeviceptr devPtr;
float* d_data;
// Allocation using driver API
cuMemAlloc(&devPtr, size);
d_data = (float*)devPtr;
// Allocation using runtime API
cudaMalloc(&d_data, size);
devPtr = (CUdeviceptr)d_data;

特别是，这意味着使用驱动程序 API 编写的应用程序可以调用使用运行时 API 编写的库（例如 cuFFT、cuBLAS 等）。

参考手册的设备和版本管理部分的所有功能都可以互换使用。

17.5. 驱动程序入口点访问

17.5.1. 简介

Driver Entry Point Access APIs提供了一种检索 CUDA 驱动程序函数地址的方法。从 CUDA 11.3 开始，用户可以使用从这些 API 获取的函数指针调用可用的 CUDA 驱动程序 API。

这些 API 提供的功能类似于它们的对应项，POSIX 平台上的 dlsym 和 Windows 上的 GetProcAddress。提供的 API 将允许用户：

· 使用CUDA Driver API.

· 使用CUDA Runtime API.

· 请求 CUDA 驱动程序函数的 每线程默认流 版本。有关更多详细信息，请参阅 检索每线程默认流版本

· 使用较新的驱动程序访问较旧工具包上的新 CUDA 功能。

17.5.2. 驱动程序函数类型定义

为了帮助检索 CUDA 驱动程序 API 入口点，CUDA 工具包提供了对包含所有 CUDA 驱动程序 API 的函数指针定义的标头的访问。这些标头与 CUDA 工具包一起安装，并在工具包的include/目录中提供。下表总结了包含typedefs每个 CUDA API 头文件的头文件。

上面的标头本身没有定义实际的函数指针;它们定义函数指针的 typedef。例如，cudaTypedefs.h具有以下驱动程序cuMemAlloc API 的类型定义：

typedef CUresult (CUDAAPI *PFN_cuMemAlloc_v3020)(CUdeviceptr_v2 *dptr, size_t bytesize);
typedef CUresult (CUDAAPI *PFN_cuMemAlloc_v2000)(CUdeviceptr_v1 *dptr, unsigned int bytesize);

CUDA 驱动程序符号具有基于版本的命名方案，其名称中带有_v*扩展名，但第一个版本除外。当特定 CUDA 驱动程序cuMemAlloc API 的签名或语义发生变化时，我们会递增相应驱动程序符号的版本号。对于驱动程序 API，第一个驱动程序符号名称为cuMemAlloc ，下一个符号名称为cuMemAlloc_v2 。CUDA 2.0 （2000） 中引入的第一个版本的 typedef 是 PFN_cuMemAlloc_v2000。CUDA 3.2 （3020） 中引入的下一个版本的 typedef 是PFN_cuMemAlloc_v3020 。

typedefs可用于更轻松地在代码中定义适当类型的函数指针：

PFN_cuMemAlloc_v3020 pfn_cuMemAlloc_v2;
PFN_cuMemAlloc_v2000 pfn_cuMemAlloc_v1;

如果用户对特定版本的 API 感兴趣，则首选上述方法。此外，标头具有发布已安装的 CUDA 工具包时可用的所有驱动程序符号的最新版本的预定义宏;这些 Typedef 没有_v*后缀。对于 CUDA 11.3 工具包， cuMemAlloc_v2是最新版本，因此我们还可以定义其函数指针，如下所示：

PFN_cuMemAlloc pfn_cuMemAlloc;

17.5.3. 驱动程序函数检索

使用驱动程序入口点访问 API 和适当的 typedef，我们可以获取指向任何 CUDA 驱动程序 API 的函数指针。

17.5.3.1. 使用驱动程序 API

驱动程序 API 需要 CUDA 版本作为参数，以获取所请求驱动程序符号的 ABI 兼容版本_v*。CUDA 驱动程序 API 具有用扩展表示的每个函数 ABI。例如，考虑以下cuStreamBeginCapture版本中的cudaTypedefs.h版本及其对应的typedefs版本：

// cuda.h
CUresult CUDAAPI cuStreamBeginCapture(CUstream hStream);
CUresult CUDAAPI cuStreamBeginCapture_v2(CUstream hStream, CUstreamCaptureMode mode);
// cudaTypedefs.h
typedef CUresult (CUDAAPI *PFN_cuStreamBeginCapture_v10000)(CUstream hStream);
typedef CUresult (CUDAAPI *PFN_cuStreamBeginCapture_v10010)(CUstream hStream, CUstreamCaptureMode mode);

从上面的typedefs代码片段中，版本后缀_v10000与_v10010指示上述 API 分别在 CUDA 10.0 和 CUDA 10.1 中引入。

#include <cudaTypedefs.h>
// Declare the entry points for cuStreamBeginCapture
PFN_cuStreamBeginCapture_v10000 pfn_cuStreamBeginCapture_v1;
PFN_cuStreamBeginCapture_v10010 pfn_cuStreamBeginCapture_v2;
// Get the function pointer to the cuStreamBeginCapture driver symbol
cuGetProcAddress("cuStreamBeginCapture", &pfn_cuStreamBeginCapture_v1, 10000, CU_GET_PROC_ADDRESS_DEFAULT, &driverStatus);
// Get the function pointer to the cuStreamBeginCapture_v2 driver symbol
cuGetProcAddress("cuStreamBeginCapture", &pfn_cuStreamBeginCapture_v2, 10010, CU_GET_PROC_ADDRESS_DEFAULT, &driverStatus);

参考上面的代码片段，要检索驱动程序cuStreamBeginCapture API版本_v1的地址，CUDA 版本参数应正好为 10.0 （10000）。同样，用于检索 API _v2版本地址的 CUDA 版本应为 10.1 （10010）。指定更高的 CUDA 版本来检索驱动程序 API 的特定版本可能并不总是可移植的。例如，在此处使用 11030 仍会返回_v2符号，但如果在 CUDA 11.3 中发布了假设_v3版本，则cuGetProcAddress API 将在与 CUDA 11.3 驱动程序配对时开始返回较新的_v3符号。由于 and 符号的 ABI _v2和_v3函数签名可能不同，因此使用_v10010用于_v2符号的定义调用_v3函数将表现出未定义的行为。

要检索给定 CUDA 工具包的驱动程序 API 的最新版本，我们还可以将 CUDA_VERSION 指定为version参数，并使用未版本化的 typedef 来定义函数指针。由于_v2是 CUDA 11.3 中驱动程序cuStreamBeginCapture API 的最新版本，因此下面的代码片段显示了检索它的不同方法。

// Assuming we are using CUDA 11.3 Toolkit
#include <cudaTypedefs.h>
// Declare the entry point
PFN_cuStreamBeginCapture pfn_cuStreamBeginCapture_latest;
// Intialize the entry point. Specifying CUDA_VERSION will give the function pointer to the
// cuStreamBeginCapture_v2 symbol since it is latest version on CUDA 11.3.
cuGetProcAddress("cuStreamBeginCapture", &pfn_cuStreamBeginCapture_latest, CUDA_VERSION, CU_GET_PROC_ADDRESS_DEFAULT, &driverStatus);

请注意，请求具有无效 CUDA 版本的驱动程序 API 将返回CUDA_ERROR_NOT_FOUND误差 。在上面的代码示例中，传入小于 10000 （CUDA 10.0） 的版本将无效。

17.5.3.2. 使用运行时 API

运行时 API 使用 CUDA 运行时版本获取所请求驱动程序符号的 ABI 兼容版本。在下面的代码片段中，所需的最低 CUDA 运行时版本是当时引入cuMemAllocAsync的 CUDA 11.2。

#include <cudaTypedefs.h>
// Declare the entry point
PFN_cuMemAllocAsync pfn_cuMemAllocAsync;
// Intialize the entry point. Assuming CUDA runtime version >= 11.2
cudaGetDriverEntryPoint("cuMemAllocAsync", &pfn_cuMemAllocAsync, cudaEnableDefault, &driverStatus);
// Call the entry point
if(driverStatus == cudaDriverEntryPointSuccess && pfn_cuMemAllocAsync) {
    pfn_cuMemAllocAsync(...);
}

17.5.3.3. 检索每线程默认流版本

某些 CUDA 驱动程序 API 可以配置为具有 默认 流或 每线程默认流 语义。具有 每线程默认流 语义的驱动程序 API 的名称中带有 后缀_ptsz 或 _ptds 。例如， cuLaunchKernel具有名为cuLaunchKernel_ptsz的 每线程默认流 变体。使用驱动程序入口点访问 API，用户可以请求驱动程序 API 的 每线程 默认流版本，而不是 默认流 版本。为 默认 流或 每线程默认流 语义配置 CUDA 驱动程序 API cuLaunchKernel会影响同步行为。更多细节可以 在这里 找到。

可以通过以下方法之一获取驱动程序 API 的默认流或 每线程默认流 版本：

· 使用--default-stream per-thread编译标志或定义宏CUDA_API_PER_THREAD_DEFAULT_STREAM以获取 每个线程的默认流 行为。

· 分别使用标志强制 默认 流CU_GET_PROC_ADDRESS_LEGACY_STREAM/cudaEnableLegacyStream或CU_GET_PROC_ADDRESS_PER_THREAD_DEFAULT_STREAM/cudaEnablePerThreadDefaultStream 每线程默认流 行为。

17.5.3.4. 访问新的 CUDA 功能

始终建议安装最新的 CUDA 工具包以访问新的 CUDA 驱动程序功能，但如果由于某种原因，用户不想更新或无法访问最新的工具包，则该 API 可用于访问新的 CUDA 功能仅具有更新的 CUDA 驱动程序。为了进行讨论，让我们假设用户在 CUDA 11.3 上，并希望使用 CUDA 12.0 驱动程序中提供的新cuFoo驱动程序 API。下面的代码片段说明了这个用例：

int main()
    // Assuming we have CUDA 12.0 driver installed.
    // Manually define the prototype as cudaTypedefs.h in CUDA 11.3 does not have the cuFoo typedef
    typedef CUresult (CUDAAPI *PFN_cuFoo)(...);
    PFN_cuFoo pfn_cuFoo = NULL;
    CUdriverProcAddressQueryResult driverStatus;
    // Get the address for cuFoo API using cuGetProcAddress. Specify CUDA version as
    // 12000 since cuFoo was introduced then or get the driver version dynamically
    // using cuDriverGetVersion
    int driverVersion;
    cuDriverGetVersion(&driverVersion);
    CUresult status = cuGetProcAddress("cuFoo", &pfn_cuFoo, driverVersion, CU_GET_PROC_ADDRESS_DEFAULT, &driverStatus);
    if (status == CUDA_SUCCESS && pfn_cuFoo) {
        pfn_cuFoo(...);
    else {
        printf("Cannot retrieve the address to cuFoo - driverStatus = %d. Check if the latest driver for CUDA 12.0 is installed.\n", driverStatus);
        assert(0);
    // rest of code here
}

17.5.4. cuGetProcAddress 的潜在影响

下面是一组与 和 的潜在问题的具体cuGetProcAddress和cudaGetDriverEntryPoint理论示例。

17.5.4.1. cuGetProcAddress 与隐式链接的含义

cuDeviceGetUuid在 CUDA 9.2 中引入。此 API 在 CUDA 11.4 中引入了较新的修订版（cuDeviceGetUuid_v2）。为了保持次要版本的兼容性，在 CUDA 12.0 之前cuDeviceGetUuid不会在 cuda.h 中进行版本碰撞cuDeviceGetUuid_v2。这意味着通过cuGetProcAddress获取指向它的函数指针来调用它可能具有不同的行为。直接使用 API 的示例：

#include <cuda.h>
CUuuid uuid;
CUdevice dev;
CUresult status;
status = cuDeviceGet(&dev, 0); // Get device 0
// handle status
status = cuDeviceGetUuid(&uuid, dev) // Get uuid of device 0

在此示例中，假设用户正在使用 CUDA 11.4 进行编译。请注意，这将执行cuDeviceGetUuid 的行为，而不是_v2版本。现在使用cuGetProcAddress以下示例：

#include <cudaTypedefs.h>
CUuuid uuid;
CUdevice dev;
CUresult status;
CUdriverProcAddressQueryResult driverStatus;
status = cuDeviceGet(&dev, 0); // Get device 0
// handle status
PFN_cuDeviceGetUuid pfn_cuDeviceGetUuid;
status = cuGetProcAddress("cuDeviceGetUuid", &pfn_cuDeviceGetUuid, CUDA_VERSION, CU_GET_PROC_ADDRESS_DEFAULT, &driverStatus);
if(CUDA_SUCCESS == status && pfn_cuDeviceGetUuid) {
    // pfn_cuDeviceGetUuid points to ???
}

在此示例中，假设用户正在使用 CUDA 11.4 进行编译。这将获得cuDeviceGetUuid_v2 的函数指针。然后，调用函数指针将调用新的_v2函数，这与前面示例中所示的cuDeviceGetUuid函数不同。

17.5.4.2. cuGetProcAddress 中的编译时间与运行时版本使用情况

让我们处理同样的问题并进行一个小调整。最后一个示例使用 CUDA_VERSION 的编译时常量来确定要获取的函数指针。如果用户使用 cuDriverGetVersion或 cudaDriverGetVersion传递给cuGetProcAddress 动态查询驱动程序版本，则会出现更多复杂情况。例：

#include <cudaTypedefs.h>
CUuuid uuid;
CUdevice dev;
CUresult status;
int cudaVersion;
CUdriverProcAddressQueryResult driverStatus;
status = cuDeviceGet(&dev, 0); // Get device 0
// handle status
status = cuDriverGetVersion(&cudaVersion);
// handle status
PFN_cuDeviceGetUuid pfn_cuDeviceGetUuid;
status = cuGetProcAddress("cuDeviceGetUuid", &pfn_cuDeviceGetUuid, cudaVersion, CU_GET_PROC_ADDRESS_DEFAULT, &driverStatus);
if(CUDA_SUCCESS == status && pfn_cuDeviceGetUuid) {
    // pfn_cuDeviceGetUuid points to ???
}

在此示例中，假设用户正在使用 CUDA 11.3 进行编译。用户将使用获取cuDeviceGetUuid（不是_v2版本）的已知行为来调试、测试和部署此应用程序。由于 CUDA 保证了次要版本之间的 ABI 兼容性，因此预计在驱动程序升级到 CUDA 11.4（无需更新工具包和运行时）后，同一应用程序将运行，而无需重新编译。不过，这将具有未定义的行为，因为现在PFN_cuDeviceGetUuid仍将是原始版本的签名，但由于cudaVersion现在是 11040 （CUDA 11.4），cuGetProcAddress将返回指向_v2版本的函数指针，这意味着调用它可能具有未定义的行为。

请注意，在这种情况下，原始（不是_v2版本）typedef如下所示：

typedef CUresult (CUDAAPI *PFN_cuDeviceGetUuid_v9020)(CUuuid *uuid, CUdevice_v1 dev);

但是_v2版本的typedef看起来像：

typedef CUresult (CUDAAPI *PFN_cuDeviceGetUuid_v11040)(CUuuid *uuid, CUdevice_v1 dev);

因此，在这种情况下，API/ABI 将是相同的，运行时 API 调用可能不会引起问题，只会导致未知 uuid 返回的可能性。在对API/ABI的影响中，我们讨论了 API/ABI 兼容性的一个更成问题的情况。

17.5.4.3. 显式版本检查的 API 版本颠簸

上面，是一个具体的例子。例如，现在让我们使用一个理论示例，该示例仍然存在驱动程序版本之间的兼容性问题。例：

CUresult cuFoo(int bar); // Introduced in CUDA 11.4
CUresult cuFoo_v2(int bar); // Introduced in CUDA 11.5
CUresult cuFoo_v3(int bar, void* jazz); // Introduced in CUDA 11.6
typedef CUresult (CUDAAPI *PFN_cuFoo_v11040)(int bar);
typedef CUresult (CUDAAPI *PFN_cuFoo_v11050)(int bar);
typedef CUresult (CUDAAPI *PFN_cuFoo_v11060)(int bar, void* jazz);

请注意，自 CUDA 11.4 中最初创建以来，API 已被修改两次，CUDA 11.6 中的最新 API 也修改了函数的 API/ABI 接口。针对 CUDA 11.5 编译的用户代码中的用法是：

#include <cuda.h>
#include <cudaTypedefs.h>
CUresult status;
int cudaVersion;
CUdriverProcAddressQueryResult driverStatus;
status = cuDriverGetVersion(&cudaVersion);
// handle status
PFN_cuFoo_v11040 pfn_cuFoo_v11040;
PFN_cuFoo_v11050 pfn_cuFoo_v11050;
if(cudaVersion < 11050 ) {
    // We know to get the CUDA 11.4 version
    status = cuGetProcAddress("cuFoo", &pfn_cuFoo_v11040, cudaVersion, CU_GET_PROC_ADDRESS_DEFAULT, &driverStatus);
    // Handle status and validating pfn_cuFoo_v11040
else {
    // Assume >= CUDA 11.5 version we can use the second version
    status = cuGetProcAddress("cuFoo", &pfn_cuFoo_v11050, cudaVersion, CU_GET_PROC_ADDRESS_DEFAULT, &driverStatus);
    // Handle status and validating pfn_cuFoo_v11050
}

在此示例中，如果没有 CUDA 11.6 中新 typedef 的更新，并且使用这些新的 typedef 和大小写处理重新编译应用程序，应用程序将返回cuFoo_v3函数指针，然后该函数的任何使用都会导致未定义的行为。此示例的重点是说明，即使是显式的版本检查cuGetProcAddress也可能无法安全地涵盖 CUDA 主要版本中的次要版本颠簸。

17.5.4.4. 运行时 API 使用问题

上述示例重点介绍了驱动程序 API 用于获取指向驱动程序 API 的函数指针的用法问题。现在，我们将讨论 的运行时 API 使用cudaApiGetDriverEntryPoint的潜在问题。

我们将首先使用类似于上述的运行时 API。

#include <cuda.h>
#include <cudaTypedefs.h>
#include <cuda_runtime.h>
CUresult status;
cudaError_t error;
int driverVersion, runtimeVersion;
CUdriverProcAddressQueryResult driverStatus;
// Ask the runtime for the function
PFN_cuDeviceGetUuid pfn_cuDeviceGetUuidRuntime;
error = cudaGetDriverEntryPoint ("cuDeviceGetUuid", &pfn_cuDeviceGetUuidRuntime, cudaEnableDefault, &driverStatus);
if(cudaSuccess == error && pfn_cuDeviceGetUuidRuntime) {
    // pfn_cuDeviceGetUuid points to ???
}

此示例中的函数指针甚至比上面的仅驱动程序示例更复杂，因为无法控制要获取哪个版本的函数;它将始终获取当前 CUDA 运行时版本的 API。有关详细信息，请参阅下表：

V11.3 => 11.3 CUDA Runtime and Toolkit (includes header files cuda.h and cudaTypedefs.h)
V11.4 => 11.4 CUDA Runtime and Toolkit (includes header files cuda.h and cudaTypedefs.h)
v1 => cuDeviceGetUuid
v2 => cuDeviceGetUuid_v2
x => Implies the typedef function pointer won't match the returned
     function pointer.  In these cases, the typedef at compile time
     using a CUDA 11.4 runtime, would match the _v2 version, but the
     returned function pointer would be the original (non _v2) function.

表中的问题来自较新的 CUDA 11.4 运行时和工具包以及较旧的驱动程序 （CUDA 11.3） 组合，在上面标记为 v1x。此组合将使驱动程序返回指向旧函数（非 _v2）的指针，但应用程序中使用的 typedef 将用于新函数指针。

17.5.4.5. 运行时 API 和动态版本控制的问题

当我们考虑编译应用程序的 CUDA 版本、CUDA 运行时版本和应用程序动态链接的 CUDA 驱动程序版本的不同组合时，会出现更多复杂性。

#include <cuda.h>
#include <cudaTypedefs.h>
#include <cuda_runtime.h>
CUresult status;
cudaError_t error;
int driverVersion, runtimeVersion;
CUdriverProcAddressQueryResult driverStatus;
enum cudaDriverEntryPointQueryResult runtimeStatus;
PFN_cuDeviceGetUuid pfn_cuDeviceGetUuidDriver;
status = cuGetProcAddress("cuDeviceGetUuid", &pfn_cuDeviceGetUuidDriver, CUDA_VERSION, CU_GET_PROC_ADDRESS_DEFAULT, &driverStatus);
if(CUDA_SUCCESS == status && pfn_cuDeviceGetUuidDriver) {
    // pfn_cuDeviceGetUuidDriver points to ???
// Ask the runtime for the function
PFN_cuDeviceGetUuid pfn_cuDeviceGetUuidRuntime;
error = cudaGetDriverEntryPoint ("cuDeviceGetUuid", &pfn_cuDeviceGetUuidRuntime, cudaEnableDefault, &runtimeStatus);
if(cudaSuccess == error && pfn_cuDeviceGetUuidRuntime) {
    // pfn_cuDeviceGetUuidRuntime points to ???
// Ask the driver for the function based on the driver version (obtained via runtime)
error = cudaDriverGetVersion(&driverVersion);