Mercurial > hg > Papers > 2020 > soto-midterm
diff src/cuLaunchKernel.cbc @ 1:73127e0ab57c
(none)
author | soto@cr.ie.u-ryukyu.ac.jp |
---|---|
date | Tue, 08 Sep 2020 18:38:08 +0900 |
parents | |
children |
line wrap: on
line diff
--- /dev/null Thu Jan 01 00:00:00 1970 +0000 +++ b/src/cuLaunchKernel.cbc Tue Sep 08 18:38:08 2020 +0900 @@ -0,0 +1,21 @@ +int computeblockDim(int count, int maxThreadPerBlock) { + return count < maxThreadPerBlock ? count : maxThreadPerBlock; +} + +__code execCUDAExecutor(struct CUDAExecutor* executor, struct Context* task, __code next(...)) { + // check data parallelism task + if (task->iterate) { + struct MultiDimIterator* iterator = &task->iterator->iterator->MultiDimIterator; + + // compute block thread size + int blockDimX = computeblockDim(iterator->x, executor->maxThreadPerBlock); + int blockDimY = computeblockDim(iterator->y, executor->maxThreadPerBlock); + int blockDimZ = computeblockDim(iterator->z, executor->maxThreadPerBlock); + + checkCudaErrors(cuLaunchKernel(task->function, + iterator->x/blockDimX, iterator->y/blockDimY, iterator->z/blockDimZ, + blockDimX, blockDimY, blockDimZ, + 0, NULL, (void**)executor->kernelParams, NULL)); + } + ... +}