根据CUDA数学APi,许多数学函数,如正弦和余弦,都在软件(函数)和硬件(内在函数)中实现.这些内在函数可能使用GPU的特殊功能单元,那么软件实现的重点是什么?这不比硬件实现慢吗?
我正试图找到我对Julia源代码的看法,即codegen.cpp.他们使用一个jl_value_t引用自身的结构:
#define JL_DATA_TYPE \
struct _jl_value_t *type;
typedef struct _jl_value_t {
JL_DATA_TYPE
} jl_value_t;
Run Code Online (Sandbox Code Playgroud)
在eclipse中调试源代码时,这似乎不包含任何有用的信息,但它经常被使用.我该如何解释这个结构?它包含哪些信息?
对于我的项目,我正在以两种不同的方式为某些功能生成PTX指令.第一种方法使用CUDA C来实现函数,使用nvcc来编译它们 nvcc -ptx <file>.cu -o <file>.ptx.另一种方法用不同的语言编写代码,从中生成LLVM IR,并使用NVPTX后端将其编译为ptx.我在这里遇到的问题是某些功能在第二种情况下表现更差.其他功能导致或多或少的可比性能.
现在我想知道为什么某些函数的性能存在这样的差异(以及为什么其他函数没有),但是使用nsight进行分析并没有给我任何好的想法.
我发现的唯一区别是寄存器使用情况.在生成的ptx代码中,我可以看到以下内容:
使用nvcc编译
.reg .u32 %r<8>;
.reg .u64 %rd<17>;
.reg .f32 %f<8>;
.reg .pred %p<5>;
Run Code Online (Sandbox Code Playgroud)
使用nvptx编译
.reg .pred %p<396>;
.reg .s16 %rc<396>;
.reg .s16 %rs<396>;
.reg .s32 %r<396>;
.reg .s64 %rl<396>;
.reg .f64 %fl<396>;
Run Code Online (Sandbox Code Playgroud)
据我所知,这表示使用的虚拟寄存器的数量和类型,但正如您可以清楚地看到的,在第二种情况下这是不正确的.在使用nsight进行性能分析后,我可以看到实际使用的寄存器/线程的数量在第一种情况下为8,在第二种情况下为31.当然,这可能表明为什么第二种情况下的代码速度较慢,但问题是我使用NVPTX从LLVM IR编译到ptx的所有函数都存在这个问题.它们都有396个使用过的虚拟寄存器和nsight报告31个所有寄存器/线程,尽管某些函数产生的性能几乎与第一种情况完全相同.
这个注册是我减速的问题吗?为什么它不影响所有功能?如果不是,可能导致经济放缓的原因是什么?你能给出我应该看的方向的任何提示吗?
谢谢!
(使用的LLVM版本是3.3)
编辑:我注意到的另一个不同之处是失速原因:
NVCC:

NVPTX:

显然,"其他"原因相对增加.也许这可以解释这个问题?
编辑:添加了ptx源代码
此处显示的函数将数据从全局内存复制到共享内存.然后每个线程将其自己的元素和前一个元素与数组中的最后一个元素进行比较.如果比较为正,则将索引写入输出数组.
1)使用NVPTX将LLVM IR编译为PTX
// .globl julia_cuda_find_weighted_median18585
.entry julia_cuda_find_weighted_median18585(
.param .u64 .ptr .global .align 4 julia_cuda_find_weighted_median18585_param_0,
.param .u64 .ptr .global .align 4 julia_cuda_find_weighted_median18585_param_1
) // @julia_cuda_find_weighted_median18585
{
.reg …Run Code Online (Sandbox Code Playgroud) 我注意到有几种方法可以在C中的其他结构中定义结构:
struct s {
int abc;
struct {
int a;
};
struct {
int b;
} intern;
struct i {
int c;
};
struct i2 {
int d;
} intern2;
struct i3 {
int e;
};
struct i3 intern3;
};
Run Code Online (Sandbox Code Playgroud)
这个结构使用gcc或g ++编译很好,所以我假设所有参数都可以通过某种方式访问.我试过这样的:
int main(int argc, char const *argv[])
{
struct s mystruct;
mystruct.abc = 0;
mystruct.a = 1;
mystruct.intern.b = 2;
mystruct.c = 3; // <-- does not compile
mystruct.intern2.d = 4;
mystruct.intern3.e = 5;
return 0;
}
Run Code Online (Sandbox Code Playgroud)
除了访问mystruct.c …
例如cudaMemcpy和cuMemcpy?我可以看到函数定义不同,但我的意思是API.为什么有api开头,cu...一个开头cuda...?应该何时使用每个API?