| 682 | |
| 683 | |
| 684 | qcomp* gpu_getCacheOfSize(qindex numElemsPerThread, qindex numThreads) { |
| 685 | #if COMPILE_CUDA |
| 686 | |
| 687 | // do not interfere with existing kernels using the cache |
| 688 | gpu_sync(); |
| 689 | |
| 690 | qindex numNewElems = numElemsPerThread * numThreads; |
| 691 | |
| 692 | // return existing cache if it's already sufficiently big |
| 693 | if (numNewElems <= gpuCacheLen) |
| 694 | return gpuCache; |
| 695 | |
| 696 | // otherwise, resize the cache |
| 697 | gpuCacheLen = numNewElems; |
| 698 | CUDA_CHECK( cudaFree(gpuCache) ); // nullptr fine to free |
| 699 | CUDA_CHECK( cudaMalloc(&gpuCache, gpuCacheLen * sizeof *gpuCache) ); |
| 700 | |
| 701 | return gpuCache; |
| 702 | |
| 703 | #else |
| 704 | error_gpuCacheModifiedButGpuNotCompiled(); |
| 705 | return nullptr; |
| 706 | #endif |
| 707 | } |
| 708 | |
| 709 | |
| 710 | void gpu_clearCache() { |
no test coverage detected