我正在尝试减少CUDA,我真的是一个新手.我目前正在研究NVIDIA的示例代码.
我想我真的不确定如何设置块大小和网格大小,特别是当我的输入数组大于(512 X 512)而不是单个块大小时.
这是代码.
template <unsigned int blockSize>
__global__ void reduce6(int *g_idata, int *g_odata, unsigned int n)
{
extern __shared__ int sdata[];
unsigned int tid = threadIdx.x;
unsigned int i = blockIdx.x*(blockSize*2) + tid;
unsigned int gridSize = blockSize*2*gridDim.x;
sdata[tid] = 0;
while (i < n)
{
sdata[tid] += g_idata[i] + g_idata[i+blockSize];
i += gridSize;
}
__syncthreads();
if (blockSize >= 512) { if (tid < 256) { sdata[tid] += sdata[tid + 256]; } __syncthreads(); }
if (blockSize …我正在尝试编写一个函数,它接受一组未分类的键/值对,例如
<7, 4>
<2, 8>
<3, 1>
<2, 2>
<1, 5>
<7, 1>
<3, 8>
<7, 2>
并按键对它们进行排序,同时使用相同的键减少对的值:
<1, 5>
<2, 10>
<3, 9>
<7, 7>
目前,我正在使用__device__类似下面的一个函数,它本质上是一个bitonic排序,它将组合相同键的值并将旧数据设置为一个无限大的值(仅99用于现在),以便后续的bitonic排序将筛选它们到底部并且数组被int *删除的值切割.
__device__ void interBitonicSortReduce(int2 *sdata, int tid, int recordNum, int *removed) {
int n = MIN(DEFAULT_DIMBLOCK, recordNum);
for (int k = 2; k <= n; k *= 2) {
for (int j = k / 2; j > 0; j /= 2) {
int ixj = tid …我刚刚开始学习CUDA编程,我对减少有些困惑.
我知道全局内存与共享内存相比有很多访问延迟,但是我可以使用全局内存来(至少)模拟类似于共享内存的行为吗？
例如,我想总结一个长度正好的大数组的元素BLOCK_SIZE * THREAD_SIZE(网格和块的维度都是幂2),我试图使用下面的代码:
__global__ void parallelSum(unsigned int* array) {
unsigned int totalThreadsNum = gridDim.x * blockDim.x;
unsigned int idx = blockDim.x * blockIdx.x + threadIdx.x;
int i = totalThreadsNum / 2;
while (i != 0) {
if (idx < i) {
array[idx] += array[idx + i];
}
__syncthreads();
i /= 2;
}
}
我比较了这段代码的结果和在主机上串行生成的结果,奇怪的是:有时结果是相同的,但有时它们显然是不同的.在这里使用全局内存是否有任何原因？
使用 cub::BlockRadixSort 在块内排序时，如果元素数量过多，我们该如何处理？如果我们将切片大小设置得太大，临时存储的共享内存很快就无法容纳它。如果我们将其拆分为多个图块，那么在对每个图块进行排序后如何对其进行后处理？