Cuda学习笔记

CUDA C简介

基本操作

以下是调用GPU的基本操作代码。代码作用是将两个数相加。
其中要注意的是:
1. cudaMemcpy() 函数前两个参数传递的是地址。
2. cudaMalloc() 函数原型为：

cudaError_t cudaMalloc (void **devPtr, size_t  size );   

所以调用时首先将第一个参数强制转换为 (void **) 类型，再取地址 & 得到之前定义的那个一维指针地址。

代码如下：

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>

__global__ void add(int a, int b, int* c){
    *c = a + b;
}

int main()
{
    int c[10];
    int* dev_c;

    cudaMalloc( (void**)&dev_c, sizeof(int) ) ;

    add<<<1, 1>>>(2, 7, dev_c);
    // 第一个参数和第二个参数都传的是地址.
    // int c;
    // cudaMemcpy( &c, dev_c, sizeof(int),cudaMemcpyDeviceToHost ) ;
    // 以下代码传址传的是数组c的首地址
    cudaMemcpy( c, dev_c, sizeof(int),cudaMemcpyDeviceToHost ) ;

    printf("%d\n", c[0]);
    cudaFree(dev_c);
    return 0;
}

读取GPU的信息

用如下方式读取GPU信息：

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>

int main()
{
    cudaDeviceProp prop;

    int count;
    cudaGetDeviceCount(&count);

    for (int i = 0; i < count; ++i)
    {
        cudaGetDeviceProperties(&prop, i);
        printf("  --- General Information for device %d --- \n", i);
        printf("Name:  %s\n", prop.name);
        printf("Compute capability: %d.%d\n", prop.major, prop.minor);
        printf("Clock rate: %d\n", prop.clockRate);
        printf("Device copy overlap:");
        if(prop.deviceOverlap)
            printf("Enable\n");
        else
            printf("Disable\n");

        printf("Kernel execition timeout:");
        if(prop.kernelExecTimeoutEnabled)
            puts("Enable");
        else
            puts("Disable");

        printf("  --- Memory Information for device %d ---\n", i);
        printf("Total global mem: %ld\n", prop.totalGlobalMem);
        printf("Total constant mem: %ld\n", prop.totalConstMem);
        printf("Max mem pitch: %ld\n", prop.memPitch);
        printf("Texture Alignment: %ld\n", prop.textureAlignment);
        printf("  --- MP Information for device %d --- \n", i);
        printf("Multiprocessor count %d\n", prop.multiProcessorCount);
        printf("Shared mem per mp : %ld\n", prop.sharedMemPerBlock);
        printf("Registers per mp: %d\n", prop.regsPerBlock);
        printf("Threads in warp: %d\n", prop.warpSize);
        printf("Max threads per block: %d\n", prop.maxThreadsPerBlock);
        printf("Max therad dimensions: (%d, %d, %d)\n",  prop.maxThreadsDim[0], prop.maxThreadsDim[1], prop.maxThreadsDim[2]);

        printf("Max grid dimensions: (%d, %d, %d)\n", prop.maxGridSize[0], prop.maxGridSize[1], prop.maxGridSize[2]);
        puts("");

    }

    return 0;
}

CUDA C并行编程

向量和

使用CUDA C并行编程来计算向量与向量的和：

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>

const int N = 10;

__global__ void add(int* a, int* b, int* c)
{
    // blockIdx - 内置变量，保存当前执行设备代码的线程块的索引
    int tid = blockIdx.x;
    if (tid < N)
    {
        c[tid] = a[tid] + b[tid];
    }
}

int main()
{
    int a[N], b[N], c[N];
    int *dev_a, *dev_b, *dev_c;

    // GPU内存分配
    cudaMalloc((void**)&dev_a, N * sizeof(int));
    cudaMalloc((void**)&dev_b, N * sizeof(int));
    cudaMalloc((void**)&dev_c, N * sizeof(int));

    // 在CPU上给a, b 赋值
    for (int i = 0; i < N; ++i)
    {
        a[i] = -i;
        b[i] = i * i;
    }

    // 将数组a, b复制到GPU
    cudaMemcpy(dev_a, a, N * sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy(dev_b, b, N * sizeof(int), cudaMemcpyHostToDevice);

    // <<<a, b>>> 第一个参数a表示设备在执行核函数时使用的并行线程块的数量
    // 以下代码创建N个线程块在GPU上运行
    add<<<N, 1>>>(dev_a, dev_b, dev_c);

    // 将结果dev_c复制到CPU
    cudaMemcpy(c, dev_c, N * sizeof(int), cudaMemcpyDeviceToHost);

    for (int i = 0; i < N; ++i)
    {
        printf("%d + %d = %d\n", a[i], b[i], c[i]);
    }

    cudaFree(dev_a);
    cudaFree(dev_b);
    cudaFree(dev_c);

    return 0;
}

Julia集

首先要在VS和系统路径中加入glut，之后头文件导入就行了。
CPU版本：

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
//#include "cuda_gl_interop.h"
#include <gl/glut.h>
#include <stdio.h>

#define DIM 1000

struct CPUBitmap
{
    unsigned char *pixels;    /*像素点的总个数*/
    int x, y;                 /*图像的长宽*/
    void *dataBlock;          /*  */

    void (*bitmapExit)(void*);  /*这是一个函数 */

    CPUBitmap( int width, int height, void *d = NULL )
    {
        pixels = new unsigned char[width * height * 4];   /*计算总的像素点数，并分配新的空间*/
        x = width;                                        /*图像的宽*/
        y = height;                                       /*图像的高*/
        dataBlock = d;                                    /* */
    }

    /*析构函数*/
    ~CPUBitmap()
    {
        /*删除像素点*/
        delete [] pixels;       
    }
    /*取得所有像素点*/       
    unsigned char* get_ptr( void ) const   { return pixels; }
    /*取得图片总大小*/
    long image_size( void ) const { return x * y * 4; }

    void display_and_exit( void(*e)(void*) = NULL )
    {
        CPUBitmap**   bitmap = get_bitmap_ptr();
        *bitmap = this;
        bitmapExit = e;

        // a bug in the Windows GLUT implementation prevents us from
        // passing zero arguments to glutInit()
        int c=1;
        char* dummy = "";

        /*glutInit,对 GLUT (OpenGl 里面的一个工具包，包含很多函数)进行初始化,这个函数必须在其它的 GLUT使用之前调用一次。其格式比较死板,一般照抄这句glutInit(&argc, argv)就可以了*/

        glutInit( &c, &dummy );        

        /*设置显示方式,其中 GLUT_RGBA 表示使用 RGBA 颜色,与之对应的还有GLUT_INDEX(表示使用索引颜色) ；GLUT_SINGLE 表示使用单缓冲,。与之对应的还有 GLUT_DOUBLE(使用双缓冲)。*/    
        glutInitDisplayMode( GLUT_SINGLE | GLUT_RGBA );

        /*这个也简单,设置窗口的大小*/

        glutInitWindowSize( x, y );

        /*根据前面设置的信息创建窗口。参数将被作为窗口的标题。注意:窗口被创建后,并不立即显示到屏幕上。需要调用 glutMainLoop 才能看到窗口。*/

        glutCreateWindow( "bitmap" );

        glutKeyboardFunc(Key);

        /* 设置一个函数,当需要进行画图时,这个函数就会被调用。*/
        glutDisplayFunc(Draw);

        /*显示窗口*/

        glutMainLoop();

    }

        // static method used for glut callbacks
    static CPUBitmap** get_bitmap_ptr( void )
    {
        static CPUBitmap   *gBitmap;
        return &gBitmap;
    }

        // static method used for glut callbacks
    static void Key(unsigned char key, int x, int y)
    {
        /* 如果按键按的是Esc按键，则退出程序。*/
        switch (key)
        {
            case 27:
            CPUBitmap*   bitmap = *(get_bitmap_ptr());
            if (bitmap->dataBlock != NULL && bitmap->bitmapExit != NULL)
                bitmap->bitmapExit( bitmap->dataBlock );
        }
     }

    // static method used for glut callbacks

    /* 画图 */
    static void Draw( void )
    {
        CPUBitmap*   bitmap = *(get_bitmap_ptr());

        /*设置背景颜色*/
        glClearColor( 0.0, 0.0, 0.0, 1.0 );

        /*清除。GL_COLOR_BUFFER_BIT 表示清除颜色*/
        glClear( GL_COLOR_BUFFER_BIT );

        glDrawPixels( bitmap->x, bitmap->y, GL_RGBA, GL_UNSIGNED_BYTE, bitmap->pixels );

        /*保证前面的 OpenGL 命令立即执行(而不是让它们在缓冲区中等待)。其作用跟 fflush(stdout)类似。*/
        glFlush();
    }
};

struct cuComplex
{
    float r;
    float i;
    cuComplex(float a, float b) : r(a), i(b) {}
    float magnitude2(void) {
  return r * r + i * i;}
    cuComplex operator * (const cuComplex& a){
        return cuComplex(r * a.r - i * a.i, i * a.r + r*a.i);
    }
    cuComplex operator+(const cuComplex& a)
    {
        return cuComplex(r + a.r, i + a.i);
    }
};

int julia(int x, int<