Ядро lauch, указывающее поток, но с размером общей памяти по умолчанию

Question

Мне нужно указать поток для запуска ядра в CUDA. Ядро использует некоторую разделяемую память, размер которой определен в коде ядра.

static const int cBlockSize = 256;

__global__ fooKernel(void* param)
{
    __shared__ uint32_t words[cBlockSize/16];
    // implementation follows, using 2 bits of shared memory per thread
}

Однако параметр размера разделяемой памяти идет перед параметром потока в выражении запуска ядра. Так как же сказать CUDA использовать размер разделяемой памяти, заданный кодом ядра, и игнорировать то, что в коде запуска?

fooKernel<<<N/cBlockSize, cBlockSize, /* What to put here? */, stream>>>(param);

Очевидно, я бы хотел избежать дублирования кода (cBlockSize/16)*sizeof(uint32_t) там снова. На самом деле выражение является более сложным.

Answer 1

Статически распределенная и динамически распределенная разделяемая память рассматривается во многих отношениях отдельно.

Если вы не собираетесь использовать динамически распределенную разделяемую память, безопасно передать значение по умолчанию, равное нулю, в качестве третьего параметра запуска ядра, независимо от каких-либо намерений, связанных с использованием статически распределенной разделяемой памяти.