问题标签 [gpu-constant-memory]

我有一个包含以下文件的 cuda 项目

头文件：

主文件：

和内核文件：

在main.cpp，我有float K1[3*3]我想复制到 GPU 常量内存。为此，我宣布__constant__ float K1_d[3*3];了Declarations.h并且在main.cpp我做到了

checkCudaErrors( cudaMemcpyToSymbol(&K1_d, &K1, 9*sizeof(float)) );

运行代码时出现以下错误：CUDA error at c:\users\xxx\test.cpp:241 code=13(cudaErrorInvalidSymbol) "cudaMemcpyToSymbol(&K1, &K1_inv, 9*sizeof(float))"

我试图复制一个变量，例如我在头文件中声明

并做到了：checkCudaErrors( cudaMemcpyToSymbol(&N, &K1_inv[0], sizeof(float)) );

我收到了与宏 chechCudaErrors 相同的错误消息。

如何将变量传输到 CUDA 中的常量内存？

我很难理解对 CUDA 常量内存的限制。

1. 为什么我们不能__constant__在运行时分配内存？为什么我需要在近乎全局范围的固定大小变量中进行编译？

2. 常量内存何时实际加载或卸载？我知道这cudaMemcpytoSymbol 是用来加载特定数组的，但是每个内核是否使用自己的常量内存分配？相关，绑定和解绑是否有成本，类似于绑定纹理的旧成本（也就是，使用纹理为每个内核调用增加了成本）？

3. 常量内存驻留在芯片的什么位置？

我主要对与 Pascal 和 Volta 相关的答案感兴趣。

问题01：

如果线程之间的数据请求频率非常高（每个线程从特定列中选择至少一个数据），那么使用常量内存中的数据存储是否比 Pascal 架构中的纹理有任何好处？

问题02：

如果warp中的每个线程都从常量内存中请求相同的指令数据，但每个线程都从不同的地址请求它（没有广播），那么性能会比纹理内存差吗？

问题03：

常量内存和纹理的缓存是否具有相同的性能？它们的尺寸是否相似？我可以在 Pascal 架构上为常量或纹理操作缓存维度吗？

背景：

内核中有一部分肯定会从纹理内存中的 2D 矩阵的特定列请求一些浮点数据，通常，每个线程都请求存储在来自前一个线程的请求数据“下方”的相邻单元格中的数据。

对于大多数经线，同一经线中的两个线程请求相同地址的频率低于 25%，但在不到 10% 的经线中可以达到 50%。

可以肯定的是，这个特定的列适合 cmem 用于 95% 的情况。

如果线程之间的数据请求频率非常高（每个线程从特定列中选择至少一个数据），那么使用常量内存中的数据存储是否比 Pascal 架构中的纹理有任何好处？

编辑：这是这个问题的拆分版本，以改善社区搜索

我是 CUDA 编程的初学者，有一个问题。

当我按值传递参数时，如下所示：

由于变量a和b被传递给内核函数add作为函数调用堆栈中的复制值，我猜需要一些内存空间来复制。

如果我是对的，那么这些参数是否复制到 GPU 或主机的主内存中的额外内存空间？

我想知道这个问题的原因是我应该将一个大结构传递给内核函数。

我也想过传递结构的指针，但似乎需要这些方式来为结构和每个成员变量调用cudamalloc 。

我有一些函数可以在常量设备内存中加载一个变量并启动一个内核函数。我注意到一个函数第一次在常量内存中加载一个变量需要 0.6 秒，但下一次在常量内存上的加载非常快（0.0008 秒）。无论哪个函数是主函数中的第一个，都会发生此行为。下面是一个示例代码：

为什么会这样？我可以避免吗？

__constant__我有一个内核多次使用少量内存，__constant__每次都需要将不同的值复制到内存中。最近，我需要使这个内核多流并发。

如何使每个流成为该__constant__内存的副本？

这个问题是解释 ptxas 的详细输出的延续，第一部分。

当我们用 编译一个内核.ptx文件ptxas -v，或者用 用一个.cu文件编译它时-ptxas-options=-v，我们会得到几行输出，例如：

（与链接到的问题中的示例相同；但名称拆解）

这个问题是关于最后一行的。来自其他内核的更多示例：

除了使用的寄存器数量之外，我们如何解释这一行的信息？具体来说：

• 是cmem常量记忆的缩写吗？
• 为什么有不同的类别，cmem即cmem[0]，，？cmem[2]cmem[14]
• smem可能代表shared memory; 它只是静态共享内存吗？
• 每种条目在什么情况下会出现在这一行？

我想将每个线程使用的一组初始化值复制到__global__内存中。我已将它们总结为一个结构。__global__但是，将其放入内存存在多个问题。首先，VS2015 告诉我__constant__这一行的“变量不支持动态初始化”： __constant__ initValsStruct d_initVals;

其次，它告诉我“在这一行中 没有合适的转换函数 from initValsStructto ：const void *cudaMemcpyToSymbol(d_initVals, &h_initVals, sizeof(initValsStruct));

这可能是一个非常基本的 C 或 CUDA 问题，但是将单个结构复制到__global__内存的最佳方法是什么？

我尝试了代码中的内容；我在 CUDA 开发论坛的某个地方找到了一个示例，其中一些__constant__内存（int1024 个元素的数组）以相同的方式初始化。

这是我访问它的方式：

我希望它能够以与将 initVals 结构作为参数提供给内核时相同的方式进行编译和运行

我有一个示例代码来说明这个问题：

当我使用 cuda-memcheck 运行此代码时，我得到：

这似乎表明我传入的数组不是常量，因此无法将其复制到常量内存中 - 是这样吗？如果是这样，我如何将作为输入提供给内核的数组复制到常量内存？