草庐IT

c++ - 内核参数的 CUDA C++ 模板化

coder 2023-11-14 原文

我正在尝试基于 bool 变量对 CUDA 内核进行模板化(如此处所示:Should I unify two similar kernels with an 'if' statement, risking performance loss?)，但我不断收到编译器错误，提示我的函数不是模板。我认为我只是遗漏了一些明显的东西，所以非常令人沮丧。

以下不起作用:

实用工具

#include "kernels.cuh"
//Utility functions

内核.cuh

    #ifndef KERNELS
    #define KERNELS
    template<bool approx>
    __global__ void kernel(...params...);
    #endif

内核.cu

template<bool approx>
__global__ void kernel(...params...)
{
    if(approx)
    {
       //Approximate calculation
    }
    else
    {
      //Exact calculation
    }
}

template __global__ void kernel<false>(...params...); //Error occurs here

主.cu

#include "kernels.cuh"
kernel<false><<<dimGrid,dimBlock>>>(...params...);

以下确实有效:

实用工具

#include "kernels.cuh"
//Utility functions

内核.cuh

#ifndef KERNELS
#define KERNELS
template<bool approx>
__global__ void kernel(...params...);
template<bool approx>
__global__ void kernel(...params...)
{
    if(approx)
    {
       //Approximate calculation
    }
    else
    {
      //Exact calculation
    }
}
#endif

主.cu

#include "kernels.cuh"
kernel<false><<<dimGrid,dimBlock>>>(...params...);

如果我把

template __global__ void kernel<false>(...params...);

kernels.cuh 末尾的行也有效。

我收到以下错误(均指上面标记的行):

kernel is not a template
invalid explicit instantiation declaration

如果有所不同，我会在一行中编译所有 .cu 文件，例如:

nvcc -O3 -arch=sm_21 -I. main.cu kernels.cu -o program

最佳答案

所有显式特化声明必须在模板实例化时可见。您的显式特化声明仅在 kernels.cu 翻译单元中可见，但在 main.cu 中不可见。

以下代码确实可以正常工作(除了在显式实例化指令中添加 __global__ 限定符)。

#include<cuda.h>
#include<cuda_runtime.h>
#include<stdio.h>
#include<conio.h>

template<bool approx>
__global__ void kernel()
{
    if(approx)
    {
        printf("True branch\n");
    }
    else
    {
        printf("False branch\n");
    }
}

template __global__ void kernel<false>();

int main(void) {
    kernel<false><<<1,1>>>();
    getch();
    return 0;
}

编辑

在 C++ 中，模板函数在遇到函数的显式实例化之前不会被编译。从这个角度来看，现在完全支持模板的 CUDA 的行为方式与 C++ 完全相同。

举个具体的例子，当编译器发现类似的东西

template<class T>
__global__ void kernel(...params...)
{
    ...
    T a;
    ...
}

它只检查函数语法，但不生成目标代码。所以，如果你像上面那样用一个模板函数编译一个文件，你将得到一个“空”的目标文件。这是合理的，因为编译器不知道分配给 a 的类型。

编译器只有在遇到函数模板的显式实例化时才会生成目标代码。这就是此时模板函数编译的工作方式，这种行为对多文件项目引入了限制:模板函数的实现(定义)必须与其声明位于同一文件中。所以，你不能把kernels.cuh中包含的接口(interface)从kernels.cu中分离出来，这是你的第一个版本的代码没有的主要原因编译。因此，您必须在使用模板的任何文件中同时包含接口(interface)和实现，即，您必须在 main.cu 中同时包含 kernels.cuh 和 kernels .cu.

由于没有显式实例化就不会生成代码，因此编译器可以容忍在项目中多次包含具有声明和定义的同一模板文件，而不会产生链接错误。

有几个关于在 C++ 中使用模板的教程。 An Idiot's Guide to C++ Templates - Part 1 ，除了令人恼火的标题外，还将为您提供有关该主题的逐步介绍。

关于c++ - 内核参数的 CUDA C++ 模板化，我们在Stack Overflow上找到一个类似的问题： https://stackoverflow.com/questions/19864920/

amp43codeprekernelc++templatescuda

有关c++ - 内核参数的 CUDA C++ 模板化的更多相关文章

  1. ruby-on-rails - 如何在 ruby​​ 中使用两个参数异步运行 exe？ - 2

    exe应该在我打开页面时运行。异步进程需要运行。有什么方法可以在ruby​​中使用两个参数异步运行exe吗？我已经尝试过ruby​​命令-system()、exec()但它正在等待过程完成。我需要用参数启动exe，无需等待进程完成是否有任何ruby​​gems会支持我的问题？ 最佳答案 您可以使用Process.spawn和Process.wait2:pid=Process.spawn'your.exe','--option'#Later...pid,status=Process.wait2pid您的程序将作为解释器的子进程执行。除

  2. ruby-on-rails - 如何优雅地重启 thin + nginx？ - 2

    我的瘦服务器配置了nginx，我的ROR应用程序正在它们上运行。在我发布代码更新时运行thinrestart会给我的应用程序带来一些停机时间。我试图弄清楚如何优雅地重启正在运行的Thin实例，但找不到好的解决方案。有没有人能做到这一点？ 最佳答案 #Restartjustthethinserverdescribedbythatconfigsudothin-C/etc/thin/mysite.ymlrestartNginx将继续运行并代理请求。如果您将Nginx设置为使用多个上游服务器，例如server{listen80;server

  3. ruby - 通过 erb 模板输出 ruby​​ 数组 - 2

    我正在使用puppet为ruby​​程序提供一组常量。我需要提供一组主机名，我的程序将对其进行迭代。在我之前使用的bash脚本中，我只是将它作为一个puppet变量hosts=>"host1,host2"我将其提供给bash脚本作为HOSTS=显然这对ruby​​不太适用——我需要它的格式hosts=["host1","host2"]自从phosts和putsmy_array.inspect提供输出["host1","host2"]我希望使用其中之一。不幸的是，我终其一生都无法弄清楚如何让它发挥作用。我尝试了以下各项:我发现某处他们指出我需要在函数调用前放置“function_”……这

  4. ruby - RSpec - 使用测试替身作为 block 参数 - 2

    我有一些Ruby代码，如下所示:Something.createdo|x|x.foo=barend我想编写一个测试，它使用double代替block参数x，这样我就可以调用:x_double.should_receive(:foo).with("whatever").这可能吗？ 最佳答案 specify'something'dox=doublex.should_receive(:foo=).with("whatever")Something.should_receive(:create).and_yield(x)#callthere

  5. ruby - 如何在 Ruby 中拆分参数字符串 Bash 样式？ - 2

    我正在为一个项目制作一个简单的shell，我希望像在Bash中一样解析参数字符串。foobar"helloworld"fooz应该变成:["foo","bar","helloworld","fooz"]等等。到目前为止，我一直在使用CSV::parse_line，将列分隔符设置为""和.compact输出。问题是我现在必须选择是要支持单引号还是双引号。CSV不支持超过一个分隔符。Python有一个名为shlex的模块:>>>shlex.split("Test'helloworld'foo")['Test','helloworld','foo']>>>shlex.split('Test"

  6. ruby - 检查方法参数的类型 - 2

    我不确定传递给方法的对象的类型是否正确。我可能会将一个字符串传递给一个只能处理整数的函数。某种运行时保证怎么样？我看不到比以下更好的选择:defsomeFixNumMangler(input)raise"wrongtype:integerrequired"unlessinput.class==FixNumother_stuffend有更好的选择吗？ 最佳答案 使用Kernel#Integer在使用之前转换输入的方法。当无法以任何合理的方式将输入转换为整数时，它将引发ArgumentError。defmy_method(number)

  7. ruby-on-rails - 在默认方法参数中使用 .reverse_merge 或 .merge - 2

    两者都可以defsetup(options={})options.reverse_merge:size=>25,:velocity=>10end和defsetup(options={}){:size=>25,:velocity=>10}.merge(options)end在方法的参数中分配默认值。问题是:哪个更好？您更愿意使用哪一个？在性能、代码可读性或其他方面有什么不同吗？编辑:我无意中添加了bang(!)...并不是要询问nobang方法与bang方法之间的区别 最佳答案 我倾向于使用reverse_merge方法:option

  8. ruby - 定义方法参数的条件 - 2

    我有一个只接受一个参数的方法:defmy_method(number)end如果使用number调用方法，我该如何引发错误？？通常，我如何定义方法参数的条件？比如我想在调用的时候报错:my_method(1) 最佳答案 您可以添加guard在函数的开头，如果参数无效则引发异常。例如:defmy_method(number)failArgumentError,"Inputshouldbegreaterthanorequalto2"ifnumbereputse.messageend#=>Inputshouldbegreaterthano

  9. ruby - rails 3 redirect_to 将参数传递给命名路由 - 2

    我没有找到太多关于如何执行此操作的信息，尽管有很多关于如何使用像这样的redirect_to将参数传递给重定向的建议:action=>'something',:controller=>'something'在我的应用程序中，我在路由文件中有以下内容match'profile'=>'User#show'我的表演Action是这样的defshow@user=User.find(params[:user])@title=@user.first_nameend重定向发生在同一个用户Controller中，就像这样defregister@title="Registration"@user=Use

  10. ruby - 字符串文字中的转义状态作为 `String#tr` 的参数 - 2

    对于作为String#tr参数的单引号字符串文字中反斜杠的转义状态，我觉得有些神秘。你能解释一下下面三个例子之间的对比吗？我特别不明白第二个。为了避免复杂化，我在这里使用了'd'，在双引号中转义时不会改变含义("\d"="d")。'\\'.tr('\\','x')#=>"x"'\\'.tr('\\d','x')#=>"\\"'\\'.tr('\\\d','x')#=>"x" 最佳答案 在tr中转义tr的第一个参数非常类似于正则表达式中的括号字符分组。您可以在表达式的开头使用^来否定匹配(替换任何不匹配的内容)并使用例如a-f来匹配一

随机推荐