项目首页¶
欢迎访问我的 Sphinx 项目！以下是本项目的章节内容。
目录
CUDA编程基本知识¶
CUDA Runtime API：函数以cuda开头。
CUDA Driver API：函数以cu开头。
延迟（Latency）¶
指令延迟¶
数据类型 |
字长（bits） |
操作 |
延迟（cycles） |
备注 |
|---|---|---|---|---|
整数 |
+ |
4 |
||
整数 |
32 |
* |
16 |
|
整数 |
/,mod |
特别昂贵 |
尽可能避免，或使用位操作替代 |
|
位操作 |
4 |
|||
比较 |
4 |
|||
max,min |
4 |
|||
类型转换 |
4 |
|||
浮点 |
+,*+ |
4 |
||
浮点 |
/ |
36 |
||
倒数 |
16 |
|||
平方根倒数 |
16 |
|||
浮点 |
平方根 |
32 |
实现：平方根倒数+倒数，对0和无穷大能正确得到结果；因此它处理一个warp花费32个时钟周期。 |
|
__log(x) |
16 |
|||
__sin(x) |
32 |
|||
__cos(x) |
32 |
|||
__exp(x) |
32 |
|||
__fdividef(x,y) |
20 |
除法 |
||
- |
- |
其他函数 |
更多时钟周期 |
因为实现为多个指令的组合 |
- |
- |
内存指令 |
4 |
包括从共享或全局内存中读写的任何指令，SM使用4个时钟周期来发射warp的一个内存指令 |
- |
- |
__syncthreads() |
4 |
同步 |
有时候，编译器不得不额外插入转换指令，从而引入额外的执行周期：
对操作数为char或short的函数操作，其操作数通常转换为int；
在单精度浮点数计算中，将双精度浮点数常量（不使用任何类型后缀定义）作为输入值；
在数学函数的双精度版本，将单精度浮点数变量作为输入参数。
后两种情况可以通过以下方式避免：
对单精度浮点数常量，使用f后缀定义，比如3.141592653f
对数学函数的单精度版本，也使用f后缀定义，比如sinf().
访存延迟¶
内存类型 |
延迟 |
|---|---|
L1/Shared Memory |
10-20 |
Global Memory |
400-600 |
性能指南¶
《CUDA中文手册》第5章