代码之家  ›  专栏  ›  技术社区  ›  Yngve Sneen Lindal

如何以一种程序员喜欢的方式使用CUDA常量内存?

  •  10
  • Yngve Sneen Lindal  · 技术社区  · 14 年前

    我正在使用CUDA框架开发一个数字处理应用程序。我有一些静态数据应该可以被所有线程访问,所以我把它放在了像这样的常量内存中:

    __device__ __constant__ CaseParams deviceCaseParams;
    

    我使用调用cudammcpytosym将这些参数从主机传输到设备:

    void copyMetaData(CaseParams* caseParams)
    {
        cudaMemcpyToSymbol("deviceCaseParams", caseParams, sizeof(CaseParams));
    }
    

    这很管用。

    无论如何,似乎(通过尝试和错误,以及从网上的阅读文章中)出于某种病态的原因,deviceCaseParams的声明和它的复制操作(对cudammcpytosym的调用)必须在 相同的文件 . 目前,我在一个.cu文件中有这两个,但我真的希望在一个.cuh文件中有参数struct,这样任何实现都可以看到它。这意味着,我也必须在头文件中拥有函数,但这会破坏链接(符号已经定义),因为这两个文件都包含这个头文件(因此,MS编译器和编译程序都编译了)。

    有人对这里的设计有什么建议吗?

    更新: 查看评论

    2 回复  |  直到 14 年前
        1
  •  7
  •   Tom    14 年前

    使用最新的CUDA(例如3.2),如果您在运行时查找符号(即通过将字符串作为第一个参数传递给 cudaMemcpyToSymbol 就像你在你的例子中一样)。

    另外,使用费米类设备,你可以随意改变内存( cudaMalloc ),复制到设备内存,然后将参数作为常量指针传递。编译器将识别您是否正在跨扭曲统一地访问数据,如果是这样,则将使用常量缓存。有关更多信息,请参阅CUDA编程指南。注意:您需要使用 -arch=sm_20 .

        2
  •  4
  •   Raffles    13 年前

    如果你使用的是前费米CUDA,你现在就会发现这个问题不仅仅适用于恒定内存,它也适用于你想要的CUDA方面的任何东西。我找到的唯一两个办法是:

    1. 在一个文件(.cu)中写入所有CUDA,或者
    2. 如果需要将代码分解成单独的文件,请将自己限制在单个.cu文件包含的头中。

    如果需要在CUDA和C/C++之间共享代码,或者在项目之间共享一些共同的代码,则选项2是唯一的选择。一开始看起来很不自然,但它解决了问题。你仍然需要构造你的代码,只是不能用一种典型的C类的方式。主要的开销是,每次进行编译时 一切 . 另一方面(我认为这可能是它这样工作的原因)是CUDA编译器可以一次访问所有源代码,这有利于优化。