问题：
当我增加在循环内部处理的数据量时CUDA kernel- 它会导致应用程序中止！
例外：
ManagedCuda.CudaException: 'ErrorLaunchFailed: 执行内核时设备发生异常。常见原因包括取消引用无效的设备指针和访问越界共享内存。
问题：
如果有人能阐明我在当前实现中遇到的限制以及导致应用程序崩溃的确切原因，我将不胜感激。
或者，我附上了一个完整的内核代码，如果有人可以说如何在没有抛出异常的情况下以这种方式重新建模它。这个想法是内核正在接受combinations然后对同一组data（在循环中）执行计算。因此，内部的循环计算应该是顺序的。内核本身的执行顺序无关紧要。是组合问题。
欢迎任何建议。
代码（短版，足以中止应用程序）：
extern "C"
{
__device__ __constant__ int arraySize;
__global__ void myKernel(
unsigned char* __restrict__ output,
const int* __restrict__ in1,
const int* __restrict__ in2,
const double* __restrict__ in3,
const unsigned char* __restrict__ in4)
{
for (int row = 0; row < arraySize; row++)
{
// looping over sequential data.
}
}
}
在上面的示例中，如果arraySize接近 50_000 则应用程序开始中止。使用相同类型的输入参数，如果我们覆盖或硬核到arraySize10_000 则代码成功完成。
代码 - 内核（完整版）
#iclude <cuda.h>
#include "cuda_runtime.h"
#include <device_launch_parameters.h>
#include <texture_fetch_functions.h>
#include <builtin_types.h>
#define _SIZE_T_DEFINED
#ifndef __CUDACC__
#define __CUDACC__
#endif
#ifndef __cplusplus
#define __cplusplus
#endif
texture<float2, 2> texref;
extern "C"
{
__device__ __constant__ int width;
__device__ __constant__ int limit;
__device__ __constant__ int arraySize;
__global__ void myKernel(
unsigned char* __restrict__ output,
const int* __restrict__ in1,
const int* __restrict__ in2,
const double* __restrict__ in3,
const unsigned char* __restrict__ in4)
{
int index = blockIdx.x * blockDim.x + threadIdx.x;
if (index >= limit)
return;
bool isTrue = false;
int varA = in1[index];
int varB = in2[index];
double calculatable = 0;
for (int row = 0; row < arraySize; row++)
{
if (isTrue)
{
int idx = width * row + varA;
if (!in4[idx])
continue;
calculatable = calculatable + in3[row];
isTrue = false;
}
else
{
int idx = width * row + varB;
if (!in4[idx])
continue;
calculatable = calculatable - in3[row];
isTrue = true;
}
}
if (calculatable >= 0) {
output[index] = 1;
}
}
}
代码 - 主机（完整版）
public static void test()
{
int N = 10_245_456; // size of an output
CudaContext cntxt = new CudaContext();
CUmodule cumodule = cntxt.LoadModule(@"kernel.ptx");
CudaKernel myKernel = new CudaKernel("myKernel", cumodule, cntxt);
myKernel.GridDimensions = (N + 255) / 256;
myKernel.BlockDimensions = Math.Min(N, 256);
// output
byte[] out_host = new byte[N]; // i.e. bool
var out_dev = new CudaDeviceVariable<byte>(out_host.Length);
// input
int[] in1_host = new int[N];
int[] in2_host = new int[N];
double[] in3_host = new double[50_000]; // change it to 10k and it's OK
byte[] in4_host = new byte[10_000_000]; // i.e. bool
var in1_dev = new CudaDeviceVariable<int>(in1_host.Length);
var in2_dev = new CudaDeviceVariable<int>(in2_host.Length);
var in3_dev = new CudaDeviceVariable<double>(in3_host.Length);
var in4_dev = new CudaDeviceVariable<byte>(in4_host.Length);
// copy input parameters
in1_dev.CopyToDevice(in1_host);
in2_dev.CopyToDevice(in2_host);
in3_dev.CopyToDevice(in3_host);
in4_dev.CopyToDevice(in4_host);
myKernel.SetConstantVariable("width", 2);
myKernel.SetConstantVariable("limit", N);
myKernel.SetConstantVariable("arraySize", in3_host.Length);
// exception is thrown here
myKernel.Run(out_dev.DevicePointer, in1_dev.DevicePointer, in2_dev.DevicePointer,in3_dev.DevicePointer, in4_dev.DevicePointer);
out_dev.CopyToHost(out_host);
}
分析
我最初的假设是我遇到了内存问题，但是，根据 VS 调试器，我500mb在主机环境中遇到了一些数据。所以我想无论我将多少数据复制到 GPU - 它都不应该超过1Gb甚至最大值11Gb。后来我注意到只有当内核内部的循环有许多数据记录要处理时才会发生崩溃。这让我相信我遇到了某种线程超时限制或类似的东西。没有确凿的证据。
系统
我的系统规格是16Gb,Ram和GeForce 1080 Ti 11Gb. 使用Cuda 9.1., 和managedCuda版本8.0.22（也尝试使用来自 master 分支的 9.x 版本）
编辑 1：26.04.2018刚刚测试了相同的逻辑，但仅在OpenCL. 代码不仅成功完成，而且执行时间比 好 1.5-5 倍CUDA，具体取决于输入参数的大小：
kernel void Test (global bool* output, global const int* in1, global const int* in2, global const double* in3, global const bool* in4, const int width, const int arraySize)
{
int index = get_global_id(0);
bool isTrue = false;
int varA = in1[index];
int varB = in2[index];
double calculatable = 0;
for (int row = 0; row < arraySize; row++)
{
if (isTrue)
{
int idx = width * row + varA;
if (!in4[idx]) {
continue;
}
calculatable = calculatable + in3[row];
isTrue = false;
}
else
{
int idx = width * row + varB;
if (!in4[idx]) {
continue;
}
calculatable = calculatable - in3[row];
isTrue = true;
}
}
if (calculatable >= 0)
{
output[index] = true;
}
}
我真的不想在这里开始OpenCL/CUDA战争。如果在我的原始实施中有什么我应该关注的CUDA- 请告诉我。
编辑：26.04.2018。在遵循评论部分的建议后，我能够在抛出异常之前将处理的数据量增加 3 倍。我能够通过切换到.ptx生成Release模式而不是Debug模式来实现这一点。Debug这种改进可能与以下Generate GPU Debug information事实Yes有关的OpenCL，但越来越接近。
对于CUDA文件生成，我正在使用VS2017 Community, CUDA 9.1project, v140 toolset, build for x64platform, post build events disabled, configuration type: utility. 代码生成设置为：compute_30,sm_30. 例如，我不确定为什么不是sm_70。我没有其他选择。