parallel-processing - restrict(amp) 是否比 CUDA 内核代码更严格？

Question

在 C++ AMP 中，内核函数或 lambda 用 restrict(amp) 标记，这对允许的 C++ 子集（在此处列出）施加了严格的限制。CUDA 是否允许内核函数中的 C 或 C++ 子集有更多自由？

Answer 1

从 Visual Studio 11 和 CUDA 4.1 开始，restrict(amp)函数比 CUDA 的类似__device__函数更具限制性。最值得注意的是，AMP 对如何使用指针有更多的限制。这是 AMP 的 DirectX11 计算基板的自然结果，它不允许HLSL（图形着色器）代码中的指针。相比之下，CUDA 的低层 IR 是PTX，它比 HLSL 更通用。

这是逐行比较：

| VS 11 AMP restrict(amp) functions     | CUDA 4.1 sm_2x __device__ functions  |
|------------------------------------------------------------------------------|
|* can only call functions that have    |* can only call functions that have   |
|  the restrict(amp) clause             |  the __device__ decoration           |
|* The function must be inlinable       |* need not be inlined                 |
|* The function can declare only        |* Class types are allowed             |
|  POD variables                        |                                      |
|* Lambda functions cannot              |* Lambdas are not supported, but      |
|  capture by reference and             |  user functors can hold pointers     |
|  cannot capture pointers              |                                      |
|* References and single-indirection    |* References and multiple-indirection |
|  pointers are supported only as       |  pointers are supported              |
|  local variables and function         |                                      |
|* No recursion                         |* Recursion OK                        |
|* No volatile variables                |* Volatile variables OK               |
|* No virtual functions                 |* Virtual functions OK                |
|* No pointers to functions             |* Pointers to functions OK            |
|* No pointers to member functions      |* Pointers to member functions OK     |
|* No pointers in structures            |* Pointers in structures OK           |
|* No pointers to pointers              |* Pointers to pointers OK             |
|* No goto statements                   |* goto statements OK                  |
|* No labeled statements                |* Labeled statements OK               |
|* No try, catch, or throw statements   |* No try, catch, or throw statements  |
|* No global variables                  |* Global __device__ variables OK      |
|* Static variables through tile_static |* Static variables through __shared__ |
|* No dynamic_cast                      |* No dynamic_cast                     |
|* No typeid operator                   |* No typeid operator                  |
|* No asm declarations                  |* asm declarations (inline PTX) OK    |
|* No varargs                           |* No varargs                          |

您可以在此处阅读有关restrict(amp)限制的更多信息。您可以在CUDA C Programming Guide的附录 D 中了解 CUDA 函数中的 C++ 支持。__device__