我有一个CUDA程序,其中内核寄存器将最大理论达到的占用限制在%50。因此,我决定使用共享内存,而不是寄存器,这些变量是块线程之间的常量,并且在整个内核运行过程中几乎是只读的。我不能在这里提供源代码;我所做的在概念上是这样的:
我最初的计划是:
__global__ void GPU_Kernel (...) {
__shared__ int sharedData[N]; //N:maximum amount that doesn't limit maximum occupancy
int r_1 = A; //except for this first initial
早上好。
我开始学习cuda编程,我正在研究性能。我在CUDA的网站上看到,要想有好的表现,我们应该考虑以下四点:
-warps / SM (系统多处理器),-blocks / SM,-Register,SM,-Shared内存,SM
因此,我要回顾一下第一件事情,根据GPU,我定义了内核的尺寸,这取决于每个SM的最大偏差和每个SM的块。我的任务是用一亿美元来衡量哪种方法更好。
我所做的是一个for循环,在这个循环中,我在每次迭代时启动一个内核,使占用最大化。例如,对于NVidia 1080 GPU,我读到:
int max_blocks = 32; //maximum number of
我正在尝试运行一个测试,其中有500个用户在180秒内以4MB大小的有效负载加速。我遇到了OutOfMemoryError。
Creating summariser <summary>
Created the tree successfully using C:/Users/ajay/jmeter/apache-jmeter-5.4/performancetests/PerformanceTest_Insert_500_4mb_180.jmx
Starting standalone test @ Wed Jul 27 17:25:10 EDT 2022 (1658957110058
当我阅读编程指南时,我感觉到共享内存总是会提高性能,但似乎没有。我有两个职能:
const int Ntimes=1;
__global__ void testgl(float *A, float *C, int numElements){
int ti = threadIdx.x;
int b0 = blockDim.x*blockIdx.x;
if (b0+ti < numElements){
for(int i=0;i<Ntimes;i++){
A[b0+ti]=A[b0+ti]*A[b0+ti]*10-2
对于一个CUDA kenel来说,保持寄存器/线程的低计数有什么好处吗?
我认为没有优势(速度或其他)。3 reg/线程的上下文切换速度与48个regs/线程一样快。不使用所有可用寄存器是没有意义的,除非你只是不想这样做。寄存器在内核之间不共享。这样做不对吗?
编辑: CUDA4.2编程指南中的 (5.2.3):
The number of registers used by a kernel can have a significant impact on the number
of resident warps. For example, for devices of c
我正在使用nvprof来测量已实现的占用率,并将其确定为
已入伙0.344031 0.344031 0.344031
但是使用占用率计算器,我发现75%。
研究结果如下:
Active Threads per Multiprocessor 1536
Active Warps per Multiprocessor 48
Active Thread Blocks per Multiprocessor 6
Occupancy of each Multiprocessor 75%
我使用33个寄存器,144个字节共享内存,256个线程/块,设备功能3.5。
编辑:
另外,我想让cla