Permítanme comenzar diciendo que he leído cuidadosamente todas las preguntas similares sobre SO:

Determinación de subprocesos por bloque y bloque por cuadrículaSubprocesos por SM, subprocesos por bloqueCUDA bloques e hilosUrdimbres y número óptimo de bloques

Mi intención es tratar de calcular dinámicamente (en lugar de valores de codificación rígida) para una biblioteca de redes neuronales de avance que estoy desarrollando.

Mis datosno es un enrejado cuadrado (una matriz), como ocurre con la mayoría de los ejemplos que he visto, son dos vectores que producen una matriz, con filas y columnas desiguales:

float x[6] {1.f, 1.f, 0.f, 1.f, 1.f, 0.f}; 
thrust::device_vector<float> in_vec( x, x+6 );
float y[9] {1.f, 1.f, 1.f, 1.f, 1.f, 1.f, 1.f, 1.f, 1.f};
thrust::device_vector<float> w_vec( y, y+9 );
thrust::device_vector<float> o_wec(9);
thrust::device_vector<float> mtx_vec( 9 * 6 );

float * i_ptr = thrust::raw_pointer_cast( in_vec.data() );
float * w_ptr = thrust::raw_pointer_cast( w_vec.data() );
float * out_ptr = thrust::raw_pointer_cast( mtx_vec.data() );

dim3 threadsPerBlock(9,6);
dim3 numBlocks(1,1);
prop_mtx<<<numBlocks,threadsPerBlock>>>( w_ptr, i_ptr, out_ptr, 6 );

y el núcleo:

__global__ void prop_mtx( float * w, float * i, float * o, int s ) 
{
    int x = blockIdx.x * blockDim.x + threadIdx.x; 
    int y = blockIdx.y * blockDim.y + threadIdx.y;
    o[y + x * s] = w[x] * i[y];
}

La razón por la que he tomado este enfoque es porque tiene sentido en el cálculo de ANN, cuando se trata de cálculos de vectores / matrices. Me gustaría mantener esto consistente, y AFAIK usando una cuadrícula 2D para los cálculos de Entrada de Peso * es razonable.

Tengo que calcular mis hilos por bloque como 2D con un número desigual de hilos en la cuadrícula.

Estoy usando una GTX 660, que tiene:

  CUDA Capability Major/Minor version number:    3.0
  Total amount of global memory:                 2047 MBytes 
  ( 5) Multiprocessors, (192) CUDA Cores/MP:     960 CUDA Cores
  Maximum Texture Dimension Size (x,y,z)         1D=(65536), 2D=(65536, 65536), 3D=(4096, 4096, 4096)
  Warp size:                                     32
  Maximum number of threads per multiprocessor:  2048
  Maximum number of threads per block:           1024
  Max dimension size of a thread block (x,y,z): (1024, 1024, 64)
  Max dimension size of a grid size    (x,y,z): (2147483647, 65535, 65535)

Estoy tratando de entender cómo puedo deducir / calcular el tamaño de la cuadrícula, los hilos por bloque y el número de bloques.

Supongamos que tengo un vector de peso de 800 artículos y un vector de entrada de 6500 artículos.

¿Esto implica que lo que realmente necesito es una cuadrícula 2D de 800,6500? Por lo que yo entiendo, ¿algo más proporcionará resultados incorrectos?

Sé que mis subprocesos máximos por bloque son 1024, pero debido a que es una cuadrícula 2D, lo más probable es que sea:

dim3 threadPerBlock(X,Y);

Debido al hecho de que mi cuadrícula no es una matriz cuadrada, ¿necesito calcular los hilos X, Y por bloque de una manera diferente?

¿O necesito deducir primero la cantidad de bloques necesarios?

Finalmente, dado que mi tamaño de urdimbre de hilo es 32,

¿El tamaño mínimo de la cuadrícula, independientemente de todos los demás parámetros? necesita ser al menos 32 o un múltiplo de 32? Lo necesitoal menos 32 hilos por bloque, o un tamaño de cuadrícula donde el número más pequeño es 32?

Cualquier pseudocódigo, o explicación de cómo debo hacer esto, sería muy apreciado.

Lo que he intentado es calcular el tamaño de mi cuadrícula 2D dividiendo mis datos entre 32 tamaños de ajuste. Luego consideré calcular los hilos de la cuadrícula utilizando los SM disponibles. Por ejemplo

800 weights / 5 SM, = 160 x's per SM
6500 inputs  / 5 SM, = 1300 y's per SM

Pero no sabía qué hacer a partir de ahí. Finalmente, consideré encontrar primero la relación peso-entrada:

6500/800 = 8.125

Implicando que usando el tamaño de cuadrícula mínimo de 32 para X, Y tendría que multiplicarse por 8.125 * 32 Por lo tanto, mi threadsPerBlock sería:

dim3 threadsPerBlock(32,260);

Eso es, por supuesto, 8320 hilos por bloque, que supera con creces los 1024 por bloque.

Entonces este es mi problema:¿Cómo no excedo los 1024 hilos por bloque, mientras mantengo el tamaño correcto de la cuadrícula de mis datos?

PD: Mi pregunta no se trata de optimizar el código, sino de comprender cómo distribuir los hilos y los datos de la cuadrícula por el dispositivo.