1

当我将 cublasIsamax 与常规内存分配器一起使用时 - 它工作正常。

int FindMaxIndex( const float* pVector, const size_t length )
{
    int result = 0;
    float* pDevVector = nullptr;

    if( CUBLAS_STATUS_SUCCESS != ::cudaMalloc( (void**)&pDevVector, length * sizeof(float) ) )
    {
        return -1;
    }
    if( CUBLAS_STATUS_SUCCESS !=  ::cudaMemcpy( pDevVector, pVector, length * (int)sizeof(float), cudaMemcpyHostToDevice) )
    {
        return -2;
    }
    ::cublasIsamax_v2( g_handle, length, pDevVector, 1, &result);

    if( nullptr != pDevVector )
    {
        ::cudaFree( pDevVector );
    }
    return result;
}

但是，如果尝试使用常量内存，它会失败并出现未知错误 N14。怎么了？复制到常量内存成功，但执行失败。

__constant__ float c_pIndex[ 255 ] = {0x00};

// the same function as GetIsMax but using CUBLAS function cublasIsamax_v2
int FindMaxIndexConst( const float* pVector, const size_t length, pfnMsg fnMsg )
{
    int result = 0;
    cudaError_t code = ::cudaMemcpyToSymbol( c_pIndex, pVector, length * sizeof(float), 0, cudaMemcpyHostToDevice );

    if( cudaSuccess != code )
    {
        const char* szMsg = ::cudaGetErrorString ( code );

        LogError3( L"[%d] [%hs] Could not allocate CUDA memory: %I64d pDevA", code, szMsg, (__int64)(length * sizeof(float)));
    }
    cublasStatus_t  status = ::cublasIsamax_v2( g_handle, length, c_pIndex, 1, &result);

    if( CUBLAS_STATUS_SUCCESS != status )
    {
        LogError2( L" [%d] Failed to execute <cublasIsamax_v2> : %I64d", status, (__int64)length );
    }

    return result;
}
4

1 回答 1

1

为什么不分配一个常规设备数组并将其传递给 CUBLAS？

__constant__数组不是普通__device__数组。在您的代码中，您正在获取数组的地址并将其传递给主机函数。主机上的阵列地址在设备上无效，反之亦然，如 CUDA 编程指南中所述。请参阅CUDA 编程指南：

取 a或变量的地址得到的地址__device__只能在设备代码中使用。设备内存中描述的通过 cudaGetSymbolAddress() 获得的 a或变量的地址只能在主机代码中使用。__shared____constant____device____constant__

至于__constant__通过设备指针访问内存，请参阅this answer了解为什么它不会被缓存。

最后，__constant__以这种方式使用内存，即使它被缓存在常量缓存中，由于访问模式的原因，它也是低效的。常量缓存针对 warp 中的线程间的统一访问进行了优化。isamax很可能在每个线程中访问不同的内存位置，因此访问将被序列化。因此，这将比统一访问慢 32 倍（并且可能比常规设备内存慢得多）。

于 2013-06-14T03:25:34.910 回答