void add(int *a, int *b, int *c) {
    int tid = 0; // 第0个CPU
    while (tid < N) {
        c[tid] = a[tid] + b[tid];
        tid += 1;
    }
}

#include "common/book.h"

#define N 10

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

    // 在GPU上分配内存
    HANDLE_ERROR(cudaMalloc((void **) &dev_a, N * sizeof(int)));
    HANDLE_ERROR(cudaMalloc((void **) &dev_b, N * sizeof(int)));
    HANDLE_ERROR(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
    HANDLE_ERROR(cudaMemcpy(dev_a, a, N * sizeof(int), cudaMemcpyHostToDevice));
    HANDLE_ERROR(cudaMemcpy(dev_b, b, N * sizeof(int), cudaMemcpyHostToDevice));

    add<<<N, 1>>>(dev_a, dev_b, dev_c);

    // 将数组c从GPU赋值到cpu
    HANDLE_ERROR(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]);
    }

    // 释放在GPU上分配的内存
    cudaFree(dev_a);
    cudaFree(dev_b);
    cudaFree(dev_c);

    return 0;
}

__global__ void add(int *a, int **b, int *c) {
    int tid = blockIdx.x; // 计算此索引位置的数据
    if (tid < N)
        c[tid] = a[tid] + b[tid];
}

int main(void) {
    CPUBitmap bitmap(DIM, DIM); // 实例化位图
    unsigned char *ptr = bitmap.get_ptr(); // 指向位图的指针
    kernel(ptr); // 传递给核函数
    bitmap.display_and_exit();
}


void kernel(unsigned char *ptr) {
    for (int y = 0; y < DIM; y++) {
        for (int x = 0; x < DIM; x++) {
            int offset = x + y * DIM;
            int juliaValue = julia(x, y); // 判断(x, y)对应的点是否属于 Julia 集
            ptr[offset * 4 + 0] = 255 * juliaValue;
            ptr[offset * 4 + 1] = 0;
            ptr[offset * 4 + 2] = 0;
            ptr[offset * 4 + 3] = 255;
        }
    }
}

// julia实现
int julia(int x, int y) {
    const float scale = 1.5;  // 图形的缩放函数

    // 以下将复平面原点移动到图像中心
    float jx = scale \* (float)(DIM / 2 - x) / (DIM / 2);
    float jy = scale \* (float)(DIM / 2 - y) / (DIM / 2);

    cuComplex c(-0.8, 0.156);  // 任意选择一个复数常量 C
    cuComplex a(jx, jy);

    int i = 0;
    for (i=0; i<200; i++){
        a = a \* a + c;
        if (a.magnitude2() > 1000)
            return 0;
    }

    return 1;
}


// 保存复数值的结构体：
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);
    }
};

#include "common/book.h"
#include "common/cpu_bitmap.h"

#define DIM 1000

int main(void) {
    CPUBitmap bitmap(DIM, DIM);
    unsigned char *dev_bitmap; // 申请设备指针

    HANDLE_ERROR(cudaMalloc((void **) &dev_bitmap, bitmap.image_size())); // 分配设备内存

    dim3 grid(DIM, DIM); // 初始化grid和block，dim3类型来自 CUDA 头文件，表示一个三位数组，当我们使用两个参数来初始化时，CUDA runtime 会自动把第三位设置为1
    kernel<<<grid, 1>>>(dev_bitmap);

    HANDLE_ERROR(cudaMemcpy(bitmap.get_ptr(), dev_bitmap, bitmap.image_size(), cudaMemcpyDeviceToHost)); // 设备内存复制到主机

    bitmap.display_and_exit();

    cudaFree(dev_bitmap);
}

// CUDA 版本的 kernel 实现：
__global__ void kernel(unsigned char *ptr) {
    int x = blockIdx.x;
    int y = blockIdx.y;
    int offset = x + y * gridDim.x; // 线性偏移
    int juliaValue = julia(x, y); // 判断(x, y)对应的点是否属于 Julia 集

    ptr[offset * 4 + 0] = 255 * juliaValue;
    ptr[offset * 4 + 1] = 0;
    ptr[offset * 4 + 2] = 0;
    ptr[offset * 4 + 3] = 255;
}


// julia实现，除了__device__之外，其余与 CPU 版本相同
__device__ int julia(int x, int y) {
    const float scale = 1.5;  // 图形的缩放函数

    // 以下将复平面原点移动到图像中心
    float jx = scale \* (float)(DIM / 2 - x) / (DIM / 2);
    float jy = scale \* (float)(DIM / 2 - y) / (DIM / 2);

    cuComplex c(-0.8, 0.156);  // 任意选择一个复数常量 C
    cuComplex a(jx, jy);

    int i = 0;
    for (i=0; i<200; i++){
        a = a \* a + c;
        if (a.magnitude2() > 1000)
            return 0;
    }

    return 1;

}


// cuComplex 结构体，也仅仅在__device__修饰符上有差别：
struct cuComplex {
    float r;
    float i;
    cuComplex(float a, float b) : r(a), i(b) {}

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

    __device__ cuComplex operator+(const cuComplex& a) {
        return cuComplex(r + a.r, i + a.i);
    }
};