【问题标题】:how to load values into extern shared array如何将值加载到外部共享数组中
【发布时间】:2016-08-01 16:45:43
【问题描述】:

当将数组从全局内存加载到共享内存时,共享内存中的变量不符合我的要求。

template<class T>
__global__ void kernel(T *t1,T *t2)
{
     int tid=threadIdx.x;
     extern __shared__ T array1[];
     extern __shared__ T array2[];
     array1[tid]=t1[tid];//copy (1)
     array2[tid]=t2[tid];//copy (2)
     __syncthreads();
}

事实证明array1[tid]=array2[tid]=t2[tid]。 交换copy(1)和copy(2)的位置时,结果为array1[tid]=array2[tid]=t1[tid]。 只有当我删除“extern”时,结果才是我想要的(array1 [tid] = t1 [tid],array2 [tid] = t2 [tid])。 谁能解释一下为什么?

【问题讨论】:

    标签: cuda thread-synchronization


    【解决方案1】:

    这个:

     extern __shared__ T array1[];
     extern __shared__ T array2[];
    

    不会像你想的那样工作。

    那些指针(array1array2)将指向同一个位置。

    如果您希望多个数组使用动态分配的共享内存,您必须按照programming guide 中给出的说明进行操作。像这样的:

     extern __shared__ T array1[];
     T * array2 = array1 + size_of_array_1;
    

    应该可以。

    并确保在内核启动参数中传递一个大小分配,该大小分配对于array1 的大小和array2 的大小来说都是足够的字节数

    【讨论】:

    • “static shared”和“extern shared”是否可能同时存在于同一个内核中,例如:“shared float array1[N]" 和 "extern shared T array2[]"?
    • 我相信你应该能够在同一个内核中同时使用静态和动态分配,你试过了吗?
    猜你喜欢
    • 1970-01-01
    • 2014-11-16
    • 1970-01-01
    • 2016-04-05
    • 1970-01-01
    • 1970-01-01
    • 2013-10-06
    • 1970-01-01
    相关资源
    最近更新 更多