1

我正在尝试为图像上的逻辑操作定义模板 CUDA 内核。代码如下所示：

#define AND 1
#define OR 2
#define XOR 3
#define SHL  4
#define SHR 5 

template<typename T, int opcode> 
__device__ inline T operation_lb(T a, T b)
{
    switch(opcode)
    {
    case AND:
        return a & b;
    case OR:
        return a | b;
    case XOR:
        return a ^ b;
    case SHL:
        return a << b;
    case SHR:
        return a >> b;
    default:
        return 0;
    }
}

//Logical Operation With A Constant
template<typename T, int channels, int opcode> 
__global__ void kernel_logical_constant(T* src, const T val, T* dst, int width, int height, int pitch)
{
    const int xIndex = blockIdx.x * blockDim.x + threadIdx.x;
    const int yIndex = blockIdx.y * blockDim.y + threadIdx.y;

    if(xIndex >= width || yIndex >= height) return;

    unsigned int tid = yIndex * pitch + (channels * xIndex);

    #pragma unroll
    for(int i=0; i<channels; i++)
        dst[tid + i] = operation_lb<T,opcode>(src[tid + i],val);
}

问题是当我实例化内核进行位移时，出现以下编译错误

错误 1 ​​错误：Ptx 程序集因错误而中止

内核瞬间是这样的：

template __global__ void kernel_logical_constant<unsigned char,1,SHL>(unsigned char*,unsigned char,unsigned char*,int,int,int);

unsigned char对于、unsigned short、 1 和 3 通道以及所有逻辑运算，还有 19 个类似的时刻。但只有位移瞬间，即SHL和SHR导致错误。当我删除这些瞬间时，代码会编译并完美运行。operation_lb如果我用设备函数中的任何其他操作替换位移，该代码也可以工作。我想知道这是否与由于内核的许多不同时刻而生成的 ptx 代码量有关。

我正在使用 CUDA 5.5、Visual Studio 2010、Windows 8 x64。编译为compute_1x, sm_1x.

任何帮助，将不胜感激。

4

1 回答 1

2

原始问题指定海报正在使用compute_20, sm_20. 这样，我无法使用此处的代码重现错误。但是，在评论中指出实际上sm_10正在使用。当我切换到编译时，sm_10我能够重现该错误。

它似乎是编译器中的一个错误。我这么说只是因为我不相信编译器应该生成汇编器无法处理的代码。但是除此之外，我不知道潜在的根本原因。我已经向 NVIDIA 提交了错误报告。

在我有限的测试中，它似乎只发生在unsigned charnot int。

作为一种可能的解决方法，对于 cc2.0 和更新的设备，-arch=sm_20在编译时指定。

于 2013-09-20T21:44:06.500 回答