Синхронизация динамического параллелизма CUDA (CUDA> 11.6) ⇐ C++

[Anonymous]

Мне нужно использовать такую ​​функцию, как Cudadevicesynchronize, чтобы ждать ядра, чтобы закончить выполнение. Однако после версии 11.6 больше невозможно использовать какую -либо форму синхронизации в функциях устройства. Как мы можем синхронизировать Subkernels в функции устройства?

Код: Выделить всё

__global__ void NNFeedForwardNormalMultiple(double* __restrict__ values, double* __restrict__ weigths, double* result, int inputsize, int outputsize) {
int idx = threadIdx.x + blockIdx.x * blockDim.x;
int outputidx = idx / outputsize;
int inputidx = idx % outputsize;

if (outputidx >= outputsize || inputidx >= inputsize) {
return;
}

atomicAdd(&result[outputidx], values[inputidx] * weigths[outputsize*outputidx + inputidx]);
}

__global__ void NNFeedForwardNormalActivate(double* __restrict__ biases, double* result, int size) {
int idx = threadIdx.x + blockIdx.x * blockDim.x;

if (idx >= size) {
return;
}

result[idx] = 1.0 / (1.0 + exp(-(result[idx] + biases[idx])));
}

__global__ void NNFeedForwardNormal(double* __restrict__ values, double* __restrict__ weigths, double* result, double* __restrict__ biases, int inputsize, int outputsize) {
int blocksize = (inputsize * outputsize + THREADS_PER_BLOCK - 1)/THREADS_PER_BLOCK;
NNFeedForwardNormalMultiple(values, weigths, result, inputsize, outputsize);
//normally cudaDeviceSynchronize() kind of function to wait for child kernel to finish;
NNFeedForwardNormalActivate(biases, result, outputsize);
}

Спасибо!

Подробнее здесь: https://stackoverflow.com/questions/794 ... -cuda-11-6