在CUDA 9.2中我有这样的事情：
#ifdef __CUDA_ARCH__
struct Context { float n[4]; } context;
#else
typedef __m128 Context;
#endif
struct A { float k[2]; };
struct B { float q[4]; };
struct FTransform : thrust::unary_function<A, B>
{
const Context context;
FTransform(Context context) : context(context){}
__device__ __host__ B operator()(const A& a) const
{
B b{{a.k[0], a.k[1], a.k[0]*context.n[0], a.k[1]*context.n[1]}};
return b;
}
};
void DoThrust(B* _bs, const Context& context, A* _as, uint32_t count)
{
thrust::device_ptr<B> bs = thrust::device_pointer_cast(_bs);
thrust::device_ptr<A> as = thrust::device_pointer_cast(_as);
FTransform fTransform(context);
auto first = thrust::make_transform_iterator(as, fTransform);
auto last = thrust::make_transform_iterator(as + count, fTransform);
thrust::copy(first, last, bs);
}
int main(int c, char **argv)
{
const uint32_t Count = 4;
Context context;
A* as;
B* bs;
cudaMalloc(&as, Count*sizeof(A));
cudaMalloc(&bs, Count*sizeof(B));
A hostAs[Count];
cudaMemcpy(as, hostAs, Count * sizeof(A), cudaMemcpyHostToDevice);
DoThrust(bs, context, as, Count);
B hostBs[Count];
cudaMemcpy(hostBs, bs, Count * sizeof(B), cudaMemcpyDeviceToHost);//crash
return 0;
}
然后，当我稍后调用标准cudaMemcpy（）调用结果时，我得到异常“遇到非法内存访问” .
如果我用非推力等效物替换推力代码，则没有错误，一切正常 . 试图复制到device_vectors等的各种组合我得到了不同的崩溃，似乎试图释放device_ptr由于某种原因 - 所以也许它出于某种原因？
==更新==
好吧这让人感到困惑似乎是由于我实际上更复杂的情况下的仿函数FTransform上下文成员变量 . 具体如下：
struct FTransform : thrust::unary_function<A, B>
{
#ifdef __CUDA_ARCH__
struct Context { float v[4]; } context;
#else
__m128 context;
#endif
...
};
所以我猜这是一个对齐问题，不知何故=>实际上它是，因为这是有效的：
#ifdef __CUDA_ARCH__
struct __align__(16) Context { float v[4]; } context;
#else
__m128 context;
#endif
1 回答
解决方案是确保如果在复制到GPU的推力仿函数成员（例如__m128 SSE类型）中使用对齐类型，则在NVCC的CPU和GPU代码构建过程中将它们定义为对齐 - 并且不会意外地假设如果一个类型看起来自然地对齐它在另一个通道中的等价物就可以了，否则很难理解可能发生的事情 .
所以例如在这样的代码中需要_ align _（16）：