微信公众号搜"智元新知"关注
微信扫一扫可直接关注哦！

使用CUDA内核获得堆栈溢出

分类：C&C++作者：编程之家
我编程的代码存在很大问题.
我不是专家,在来这里之前我问了很多人.也纠正了很多事情.所以,我想我已准备好向您展示代码并向您提问我的问题.
我会把整个代码放在这里,以便让你很好地理解我的问题.
我想做的事情是,如果ARRAY_SIZE对于THREAD_SIZE来说太大了,那么我将大数组的数据放入一个较小的数组中,特别是用THREAD_SIZE大小创建.
然后,我将它发送到内核并做我必须做的任何事情.但是我有问题 
isub_matrix[x*THREAD_SIZE+y]=big_matrix[x*ARRAY_SIZE+y];

由于堆栈溢出,代码停止的位置.首先,我制作了big_matrix的双指针.但freenode irc网络#cuda频道的人告诉我,cpu内存太大而无法处理它,我应该创建一个线性指针.我做到了,但我仍然有同样的堆栈溢出问题.所以,在这里…经过一些更改后更新,但还没有工作(堆栈溢出停止,但是链接和清单更新失败)

#define ARRAY_SIZE 2048
#define THREAD_SIZE 32
#define PI 3.14


int main(int argc,char** argv) 
{
        int array_plus=0,x,y;
        float time;
        //unsigned int memsize=sizeof(float)*THREAD_SIZE*THREAD_SIZE;
        //bool array_rest;
        cudaEvent_t start,stop;
        float *d_isub_matrix;

    float *big_matrix = new float[ARRAY_SIZE*ARRAY_SIZE];
    float *big_matrix2 = new float[ARRAY_SIZE*ARRAY_SIZE];
    float *isub_matrix = new float[THREAD_SIZE*THREAD_SIZE];
    float *osub_matrix = new float[THREAD_SIZE*THREAD_SIZE];

        //if the array's size is not compatible with the thread's size,it won't work.

        //array_rest=(ARRAY_SIZE*ARRAY_SIZE)/(THREAD_SIZE*THREAD_SIZE);
        //isub_matrix=(float*) malloc(memsize);
        //osub_matrix=(float*) malloc(memsize);

        if(((ARRAY_SIZE*ARRAY_SIZE)%(THREAD_SIZE*THREAD_SIZE)==0))
        {

            //allocating space in cpu memory and GPU memory for the big matrix and its sub matrixes
            //it has to be like this (lots of loops)



            //populating the big array
            for(x=0;x<ARRAY_SIZE;x++)
            {
                for(y=0;y<ARRAY_SIZE;y++)
                    big_matrix[x*ARRAY_SIZE+y]=rand()%10000;
            }

            //kind of loop for the big array

            //Start counting the time of processing (everything)
            cudaEventCreate(&start);
            cudaEventCreate(&stop);

            cudaEventRecord(start,0);

            while(array_plus<ARRAY_SIZE)
            {

                //putting the big array's values into the sub-matrix

                for(x=0;x<THREAD_SIZE;x++)
                {
                    for(y=0;y<THREAD_SIZE;y++)
                        isub_matrix[x*THREAD_SIZE+y]=big_matrix[(x+array_plus)*ARRAY_SIZE+y];
                }

                cudamalloc((void**)&d_isub_matrix,THREAD_SIZE*THREAD_SIZE*sizeof(float));
            cudamalloc((void**)&osub_matrix,THREAD_SIZE*THREAD_SIZE*sizeof(float));
            cudamemcpy(d_isub_matrix,isub_matrix,((THREAD_SIZE*THREAD_SIZE)*sizeof(float)),cudamemcpyHostToDevice);

                //call the cuda kernel

                twiddle_factor<<<1,256>>>(isub_matrix,osub_matrix);//<----

                cudamemcpy(osub_matrix,cudamemcpyDevicetoHost);

                array_plus=array_plus+THREAD_SIZE;
                for(x=0;x<THREAD_SIZE;x++)
                {
                    for(y=0;y<THREAD_SIZE;y++)
                        big_matrix2[x*THREAD_SIZE+array_plus+y]=osub_matrix[x*THREAD_SIZE+y];
                }

                array_rest=array_plus+(ARRAY_SIZE);

                cudaFree(isub_matrix);
                cudaFree(osub_matrix);
                system("PAUSE");
            }

            //Stop the time

            cudaEventRecord(stop,0);
            cudaEventSynchronize(stop);
            cudaEventelapsedtime(&time,start,stop);

            //Free memory in GPU




            printf("The processing time took... %fms to finish",time);
                    system("PAUSE");

        }
        printf("The processing time took...NAO ENTROU!");
        system("PAUSE");
        return 0;
}

//things to do: TRANSPOSITION!!!!

另一个问题是关于平行部分.
编译器(Visual Studio)说我一次搞了太多pow()和exp().
我该如何解决这个问题？

if((xIndex<THREAD_SIZE)&&(yIndex<THREAD_SIZE))
    {
        block[xIndex][yIndex]=exp(sum_sin[xIndex][yIndex])+exp(sum_cos[xIndex][yIndex]);
    }

原始代码在这里.我评论它是因为我想知道至少我的代码是否在GPU中占据了一些价值.但它甚至没有启动内核……太可悲了)

__global__ void twiddle_factor(float *isub_matrix,float *osub_matrix)
{
    __shared__ float block[THREAD_SIZE][THREAD_SIZE];
    // int x,y,z;
    unsigned int xIndex = threadIdx.x;
    unsigned int yIndex = threadIdx.y;
    /*
    int sum_sines=0.0;
    int sum_cosines=0.0;
    float sum_sin[THREAD_SIZE],sum_cos[THREAD_SIZE];
    float angle=(2*PI)/THREAD_SIZE;

    //put into shared memory the FFT calculation (F(u))

    for(x=0;x<THREAD_SIZE;x++)
    {
        for(y=0;y<THREAD_SIZE;y++)
        {
            for(z=0;z<THREAD_SIZE;z++)
            {
                sum_sines=sum_sin+sin(isub_matrix[y*THREAD_SIZE+z]*(angle*z));
                sum_cosines=sum_cos+cos(isub_matrix[y*THREAD_SIZE+z]*(angle*z));

            }
            sum_sin[x][y]=sum_sines/THREAD_SIZE;
            sum_cos[x][y]=sum_cosines/THREAD_SIZE;

        }
    }
    */

    if((xIndex<THREAD_SIZE)&&(yIndex<THREAD_SIZE))
        block[xIndex][yIndex]=pow(THREAD_SIZE,0.5);

        //block[xIndex][yIndex]=pow(exp(sum_sin[xIndex*THREAD_SIZE+yIndex])+exp(sum_cos[xIndex*THREAD_SIZE+yIndex]),0.5);

        __syncthreads();

    //transposition X x Y
    //transfer back the results into another sub-matrix that is allocated in cpu

    if((xIndex<THREAD_SIZE)&&(yIndex<THREAD_SIZE))
            osub_matrix[yIndex*THREAD_SIZE+xIndex]=block[xIndex][yIndex];



    __syncthreads();
}

感谢您阅读所有内容！

以下是整个代码：

#include <stdlib.h>
#include <stdio.h>
#include <string.h>
#include <math.h>

#define ARRAY_SIZE 2048
#define THREAD_SIZE 32
#define PI 3.14



__global__ void twiddle_factor(float *isub_matrix,float *osub_matrix)
{
    __shared__ float block[THREAD_SIZE][THREAD_SIZE];
    int x,z;
    unsigned int xIndex = threadIdx.x;
    unsigned int yIndex = threadIdx.y;

    float sum_sines=0.0;
    //float expo_sums;
    float sum_cosines=0.0;
    float sum_sin[THREAD_SIZE][THREAD_SIZE],sum_cos[THREAD_SIZE][THREAD_SIZE];
    float angle=(2*PI)/THREAD_SIZE;

    //put into shared memory the FFT calculation (F(u))

    for(x=0;x<THREAD_SIZE;x++)
    {
        for(y=0;y<THREAD_SIZE;y++)
        {
            for(z=0;z<THREAD_SIZE;z++)
            {
                sum_sines=sum_sines+sin(isub_matrix[y*THREAD_SIZE+z]*(angle*z));
                sum_cosines=sum_cosines+cos(isub_matrix[y*THREAD_SIZE+z]*(angle*z));

            }
            sum_sin[x][y]=sum_sines/THREAD_SIZE;
            sum_cos[x][y]=sum_cosines/THREAD_SIZE;

        }
    }


    if((xIndex<THREAD_SIZE)&&(yIndex<THREAD_SIZE))
    {
        block[xIndex][yIndex]=exp(sum_sin[xIndex][yIndex])+exp(sum_cos[xIndex][yIndex]);
    }




        __syncthreads();

    //transposition X x Y
    //transfer back the results into another sub-matrix that is allocated in cpu

    if((xIndex<THREAD_SIZE)&&(yIndex<THREAD_SIZE))
            osub_matrix[yIndex*THREAD_SIZE+xIndex]=block[xIndex][yIndex];



    __syncthreads();
}


int main(int argc,stop;
        float *d_isub_matrix,*d_osub_matrix;

        float *big_matrix = new float[ARRAY_SIZE*ARRAY_SIZE];
        float *big_matrix2 = new float[ARRAY_SIZE*ARRAY_SIZE];
        float *isub_matrix = new float[THREAD_SIZE*THREAD_SIZE];
        float *osub_matrix = new float[THREAD_SIZE*THREAD_SIZE];

        //if the array's size is not compatible with the thread's size,it won't work.

        //array_rest=(ARRAY_SIZE*ARRAY_SIZE)/(THREAD_SIZE*THREAD_SIZE);
        //isub_matrix=(float*) malloc(memsize);
        //osub_matrix=(float*) malloc(memsize);

        if(((ARRAY_SIZE*ARRAY_SIZE)%(THREAD_SIZE*THREAD_SIZE)==0)&&(ARRAY_SIZE>=THREAD_SIZE))
        {

            //allocating space in cpu memory and GPU memory for the big matrix and its sub matrixes
            //it has to be like this (lots of loops)



            //populating the big array
            for(x=0;x<ARRAY_SIZE;x++)
            {
                for(y=0;y<ARRAY_SIZE;y++)
                    big_matrix[x*ARRAY_SIZE+y]=rand()%10000;
            }

            //kind of loop for the big array

            //Start counting the time of processing (everything)
            cudaEventCreate(&start);
            cudaEventCreate(&stop);

            cudaEventRecord(start,0);

            while(array_plus<ARRAY_SIZE)
            {

                //putting the big array's values into the sub-matrix

                for(x=0;x<THREAD_SIZE;x++)
                {
                    for(y=0;y<THREAD_SIZE;y++)
                        isub_matrix[x*THREAD_SIZE+y]=big_matrix[x*ARRAY_SIZE+y];
                }

                cudamalloc((void**)&d_isub_matrix,THREAD_SIZE*THREAD_SIZE*sizeof(float));
                cudamalloc((void**)&d_osub_matrix,THREAD_SIZE*THREAD_SIZE*sizeof(float));
                cudamemcpy(d_isub_matrix,256>>>(d_isub_matrix,d_osub_matrix);//<----

                cudamemcpy(osub_matrix,d_osub_matrix,cudamemcpyDevicetoHost);

                array_plus=array_plus+THREAD_SIZE;
                for(x=0;x<THREAD_SIZE;x++)
                {
                    for(y=0;y<THREAD_SIZE;y++)
                        big_matrix2[x*THREAD_SIZE+array_plus+y]=osub_matrix[x*THREAD_SIZE+y];
                }


                cudaFree(isub_matrix);
                cudaFree(osub_matrix);
                cudaFree(d_osub_matrix);
                cudaFree(d_isub_matrix);
            }

            //Stop the time

            cudaEventRecord(stop,stop);

            //Free memory in GPU

解决方法

我在这段代码中看到了很多问题.

>在将数据从big_matrix复制到isub_matrix之前,您没有为isub_matrix分配内存

for(x=0;x<THREAD_SIZE;x++)
    {
        for(y=0;y<THREAD_SIZE;y++)
            isub_matrix[x*THREAD_SIZE+y]=big_matrix[x*ARRAY_SIZE+y];
    }

>您没有为isub_matrix从主机到设备执行任何cudamemcpy.在设备上为isub_matrix分配内存后,您需要复制数据.
>我在while循环中看到你正在计算相同的数据.

//putting the big array's values into the sub-matrix

        for(x=0;x<THREAD_SIZE;x++)
        {
            for(y=0;y<THREAD_SIZE;y++)
                isub_matrix[x*THREAD_SIZE+y]=big_matrix[x*ARRAY_SIZE+y];
        }

for循环应该依赖于array_plus.

我建议你这样做

for(x=0;x<THREAD_SIZE;x++)
            {
                for(y=0;y<THREAD_SIZE;y++)
                    isub_matrix[x*THREAD_SIZE+y]=big_matrix[(x+array_plus)*ARRAY_SIZE+y];
            }

>而且,我不觉得使用array_rest.那用的是什么？

基于更新版本：

我看到的错误是

>您正在使用osub_matrix作为主机和设备指针.我建议你创建另一个浮点指针并将其用于设备

float *d_osub_matrix;

cudamalloc((void**)&d_osub_matrix,THREAD_SIZE*THREAD_SIZE*sizeof(float));

并打电话.

twiddle_factor<<<1,d_osub_matrix);

然后做

cudamemcpy(osub_matrix,cudamemcpyDevicetoHost);

>顺便说一下,事实并非如此

twiddle_factor<<< 1256>>>(isub_matrix,osub_matrix);

它应该是

twiddle_factor<<< 1256>>>(d_isub_matrix,osub_matrix);

最终和完成的代码：

int main(int argc,char** argv)
{
        int array_plus=0,y;
        int array_plus_x,array_plus_y;
        float time;
        //unsigned int memsize=sizeof(float)*THREAD_SIZE*THREAD_SIZE;
        //bool array_rest;
        cudaEvent_t start,0);
            for(array_plus_x = 0; array_plus_x < ARRAY_SIZE; array_plus_x += THREAD_SIZE)
            for(array_plus_y = 0; array_plus_y < ARRAY_SIZE; array_plus_y += THREAD_SIZE)
            {


                //putting the big array's values into the sub-matrix

                for(x=0;x<THREAD_SIZE;x++)
                {
                    for(y=0;y<THREAD_SIZE;y++)
                        isub_matrix[x*THREAD_SIZE+y]=big_matrix[(x+array_plus_x)*ARRAY_SIZE+(y+array_plus_y)];
                }

                cudamalloc((void**)&d_isub_matrix,cudamemcpyHostToDevice);

                //call the cuda kernel

                dim3 block(32,32);
                twiddle_factor<<<1,block>>>(d_isub_matrix,cudamemcpyDevicetoHost);

                for(x=0;x<THREAD_SIZE;x++)
                {
                    for(y=0;y<THREAD_SIZE;y++)
                        big_matrix2[(x+array_plus_x)*ARRAY_SIZE+(y+array_plus_y)]=osub_matrix[x*THREAD_SIZE+y];
                }

                cudaFree(d_osub_matrix);
                cudaFree(d_isub_matrix);
            }

            //Stop the time

            cudaEventRecord(stop,stop);

            //Free memory in GPU

原文地址：https://www.jb51.cc/c/118224.html

版权声明：本文内容由互联网用户自发贡献，该文观点与技术仅代表作者本人。本站仅提供信息存储空间服务，不拥有所有权，不承担相关法律责任。如发现本站有涉嫌侵权/违法违规的内容， 请发送邮件至 dio@foxmail.com 举报，一经查实，本站将立刻删除。

相关推荐

宿舍管理系统——单链表+结构体实现入住、退房和查询功能（C语言版）
本程序的编译和运行环境如下（如果有运行方面的问题欢迎在评论区留言，也欢迎直接加QQ：2961439733，备注博客园或CSDN即可）： 编辑工具：Dev-C++（版本：5.11.0.0） 编译器：TDM-GCC&#160;4.9.2&#160;64-bit Release 代码生成语言标准：ISO
使用Windows API 编写按行分配的计算矩阵向量相乘的并行程序
水了一学期的院选修，万万没想到期末考试还有比较硬核的编程题，居然还涉及到多线程了，而且是调用Windows接口函数写多线程。虽然学汇编的时候有所了解，但毕竟没尝试过，思来想去，写出了下面的蹩脚代码，不足之处还望各位批评指正。题目描述如下： 采用windows接口函数，编写一个计算矩阵向量相乘的并行程
C语言实现一个走迷宫小游戏（深度优先算法）
补充一下，先前文章末尾给出的下载链接的完整代码含有部分C++的语法（使用Dev-C++并且文件扩展名为.cpp的没有影响），如果有的朋友使用的语言标准是VC6的话可能不支持，所以在修改过后再上传一版，这次直接放在文章末尾了，复制粘贴就行，希望对您有所帮助。 接上一篇万年历博文，还是那位朋友的练习题。
模N取余法实现大整数进制转换 ——C语言版（2-16进制均可）
思路如标题所说采用模N取余法，难点是这个除法过程如何实现。 个人推荐先到这篇博客学习一下，大佬的思路就是不一样：大数除法——超详细讲解 我所做的就是在上面博文代码的基础上增加了循环，用一个字符数组逆序存储余数，最后再逆序输出就得到了正确结果。具体代码及注释如下，目前我本人测试无误，能够正确运行。如果
C语言输出格式工整的日历——2乘6样式（详见本文）
本篇博客有更新！！！更新后效果图如下： 文章末尾的完整代码如不能在Dev-C++上完好运行，出现如下问题： E:\Dev-Cpp\源代码\万年历.c [Error] &#39;for&#39; loop initial declarations are only allowed in C99 or
通过调试来理解形参与实参的区别
刚开始学习模块化程序设计时，估计大家都被形参和实参搞迷糊过，尤其是遇到形参名和实参名一样时，更加晕头转向，出现一种“是谁把值传给了我，而我又传给了谁”的疑惑。我也有过类似的迷茫，更是被其他初学的同学问过很多次。想了好久还是决定整理一下，把它写成博客供更多人学习，而且是通过调试这一种直观的方式。下面进
C# Task 实现任务超时取消、超时取消然后重试 超过重试最大次数就结束。
任务超时取消 示例 public static async Task TimeoutCancelTask() { CancellationTokenSource cts = new CancellationTokenSource();//取消令牌 Task task = DoAction(cts);
Task 笔记
### 1.计时器类Stopwatch ``` Stopwatch stopwatch=new Stopwatch() stopwatch.Start();//开始计时 stopwatch.Stop();//停止计时 stopwatch.Reset();//重置计时时间 stopwatch.Rest
MindSpore自定义算子中的张量维度问题
当我们使用GPU进行快速运算时，虽然可以用成熟的深度学习框架如MindSpore和PyTorch等进行实现，但其实从速度上来说，最快不过直接使用C/C++的CUDA来实现。也正是因为如此，在MindSpore框架中支持了对CUDA实现的算子的直接调用，只是在格式规范上有一定的要求。本文主要介绍Min
CUDA指针数组Kernel函数
继上一篇文章学习使用C++存储一个不规则二维数组之后，这里介绍如何在C语言版的CUDA中实现一个不规则的二维数组。总体的实现思路跟前面一篇文章一样，使用了一个二维的指针数组来存储。其中主要的不同点大概就是在Host和Device之间的内存交互上，需要不断的分配、拷贝和释放内存，最终我们还是用一个CU
苹果市值2025年有望达4万亿美元
pythonJavaScriptjavaHTMLPHPreactjsC#AndroidCSSNode.jssqlrpython-3.xMysqLjQueryc++pandasFlutterangularIOSdjangolinuxswifttypescript路由器JSON路由器设置无线路由器h3c华三华三路由器设置华三路由器电脑软件教程arraysdocker软件图文教程Cvue.jslaravelspring-boot