CUDA学习(七十三)

简介:

C ++ 11特性:
主机编译器默认启用的C ++ 11功能也受nvcc支持,并受到本文档中所述的限制。 另外,使用-std = c ++ 11标志调用nvcc会打开所有C ++ 11功能,并且还会调用具有相应C ++ 11方言选项的主机预处理器,编译器和链接器。
Lambda表达式:
与lambda表达式关联的闭包类的所有成员函数12的执行空间说明符由编译器派生,如下所示。 如C ++ 11标准中所述,编译器在包含lambda表达式的最小块作用域,类作用域或命名空间作用域中创建闭包类型。 计算包含闭包类型的最内层函数范围,并将相应函数的执行空间说明符分配给闭包类成员函数。 如果没有封闭函数作用域,则执行空间说明符是__host__
lambda表达式和计算执行空间说明符的示例如下所示(在注释中)。

auto globalVar = [] { return 0; }; // __host__
void f1(void) {
    auto l1 = [] { return 1; }; // __host__
}
__device__ void f2(void) {
    auto l2 = [] { return 2; }; // __device__
}
__host__ __device__ void f3(void) {
    auto l3 = [] { return 3; }; // __host__ __device__
}
__device__ void f4(int(*fp)() = [] { return 4; } /* __host__ */) {
}
__global__ void f5(void) {
    auto l5 = [] { return 5; }; // __device__
}
__device__ void f6(void) {
    struct S1_t {
        static void helper(int(*fp)() = [] {return 6; } /* __device__ */) {
        }
    };
}

除非在__device____global__函数中定义lambda,否则不能在__global__函数模板实例的类型或非类型参数中使用lambda表达式的闭包类型。
例子:

template <typename T>
__global__ void foo(T in) { };
template <typename T>
struct S1_t { };
void bar(void) {
    auto temp1 = [] {};
    foo << <1, 1 >> >(temp1); // error: lambda closure type used in
                              // template type argument
    foo << <1, 1 >> >(S1_t<decltype(temp1)>()); // error: lambda closure type used in
                                                // template type argument
}

std::initializer_list:
默认情况下,CUDA编译器会隐式地考虑std :: initializer_list的成员函数,使其具有__host__ __device__执行空间说明符,因此可以直接从设备代码调用它们。 nvcc标志--nohost-device-initializer-list将禁用此行为; std :: initializer_list的成员函数将被视为__host__函数,不会直接从设备代码中调用。
例子:

#include <initializer_list>
__device__ int foo(std::initializer_list<int> in);
__device__ void bar(void)
{
    foo({ 4,5,6 }); // (a) initializer list containing only
                    // constant expressions.
    int i = 4;
    foo({ i,5,6 }); // (b) initializer list with at least one
                    // non-constant element.
                    // This form may have better performance than (a).
}

右值引用:
默认情况下,CUDA编译器会隐式考虑std :: move和std :: forward函数模板,使其具有__host__ __device__执行空间说明符,因此可以直接从设备代码调用它们。 nvcc标志 - no-hostdevice-move-forward将禁用此行为; std :: move和std :: forward将被视为__host__函数,并且不会直接从设备代码调用。\
Constexpr功能和功能模板:
默认情况下,不能从具有不兼容的执行空间13的函数调用constexpr函数。实验nvcc标志--expt-relaxed-constexpr删除此限制。 当这个标志被指定时,主机代码可以调用__device__ constexpr函数,设备代码可以调用__host__ constexpr函数。 当指定--expt-relaxedconstexpr时,nvcc将定义宏__CUDACC_RELAXED_CONSTEXPR__。 请注意,即使相应的模板标有关键字constexpr(C ++ 11 Standard Section [dcl.constexpr.p6]),函数模板实例化也可能不是constexpr函数。
timg

目录
相关文章
|
并行计算 API 调度
CUDA学习(八十八)
3.虽然__syncthreads()一直被记录为同步线程块中的所有线程,但Pascal和以前的体系结构只能在warp级别强制执行同步。 在某些情况下,只要每条经线中至少有一条线达到屏障,就可以在不被每条线执行的情况下成功实现屏障。
1691 0
|
并行计算 前端开发 Windows
|
并行计算 编译器
|
并行计算 编译器 存储
|
移动开发 并行计算 编译器
|
并行计算 C语言 编译器
|
并行计算 API C语言
|
并行计算 API 编译器
CUDA学习(六十五)
很早之前就发现云栖社区的编辑器有一个Bug,往草稿箱存博客,当草稿箱博客数超过十篇时,无法再选择十篇前的博客进行编辑
2363 0
|
存储 并行计算 API