Когда я увеличиваю развертывание с 8 до 9 циклов в моем ядре, оно прерывается с ошибкой out of resources.
Я прочитал в Как диагностировать сбой запуска CUDA из-за нехватки ресурсов? , что несоответствие параметров и чрезмерное использование регистров могут быть проблемой, но здесь это не так. .
Мое ядро вычисляет расстояние между n точками и m центроидами и выбирает для каждой точки ближайший центроид. Это работает для 8 измерений, но не для 9. Когда я устанавливаю dimensions=9 и раскомментирую две строки для вычисления расстояния, я получаю pycuda._driver.LaunchError: cuLaunchGrid failed: launch out of resources.
Как вы думаете, могло ли это быть причиной такого поведения? Какие еще иусы могут вызвать out of resources*?
Я использую Quadro FX580. Вот минимальный (ish) пример. Для раскрутки в реальном коде я использую шаблоны.
import numpy as np
from pycuda import driver, compiler, gpuarray, tools
import pycuda.autoinit
## preference
np.random.seed(20)
points = 512
dimensions = 8
nclusters = 1
## init data
data = np.random.randn(points,dimensions).astype(np.float32)
clusters = data[:nclusters]
## init cuda
kernel_code = """
// the kernel definition
__device__ __constant__ float centroids[16384];
__global__ void kmeans_kernel(float *idata,float *g_centroids,
int * cluster, float *min_dist, int numClusters, int numDim) {
int valindex = blockIdx.x * blockDim.x + threadIdx.x ;
float increased_distance,distance, minDistance;
minDistance = 10000000 ;
int nearestCentroid = 0;
for(int k=0;k<numClusters;k++){
distance = 0.0;
increased_distance = idata[valindex*numDim] -centroids[k*numDim];
distance = distance +(increased_distance * increased_distance);
increased_distance = idata[valindex*numDim+1] -centroids[k*numDim+1];
distance = distance +(increased_distance * increased_distance);
increased_distance = idata[valindex*numDim+2] -centroids[k*numDim+2];
distance = distance +(increased_distance * increased_distance);
increased_distance = idata[valindex*numDim+3] -centroids[k*numDim+3];
distance = distance +(increased_distance * increased_distance);
increased_distance = idata[valindex*numDim+4] -centroids[k*numDim+4];
distance = distance +(increased_distance * increased_distance);
increased_distance = idata[valindex*numDim+5] -centroids[k*numDim+5];
distance = distance +(increased_distance * increased_distance);
increased_distance = idata[valindex*numDim+6] -centroids[k*numDim+6];
distance = distance +(increased_distance * increased_distance);
increased_distance = idata[valindex*numDim+7] -centroids[k*numDim+7];
distance = distance +(increased_distance * increased_distance);
//increased_distance = idata[valindex*numDim+8] -centroids[k*numDim+8];
//distance = distance +(increased_distance * increased_distance);
if(distance <minDistance) {
minDistance = distance ;
nearestCentroid = k;
}
}
cluster[valindex]=nearestCentroid;
min_dist[valindex]=sqrt(minDistance);
}
"""
mod = compiler.SourceModule(kernel_code)
centroids_adrs = mod.get_global('centroids')[0]
kmeans_kernel = mod.get_function("kmeans_kernel")
clusters_gpu = gpuarray.to_gpu(clusters)
cluster = gpuarray.zeros(points, dtype=np.int32)
min_dist = gpuarray.zeros(points, dtype=np.float32)
driver.memcpy_htod(centroids_adrs,clusters)
distortion = gpuarray.zeros(points, dtype=np.float32)
block_size= 512
## start kernel
kmeans_kernel(
driver.In(data),driver.In(clusters),cluster,min_dist,
np.int32(nclusters),np.int32(dimensions),
grid = (points/block_size,1),
block = (block_size, 1, 1),
)
print cluster
print min_dist