cuda多线程间通信

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

#include <stdio.h>
#include <time.h>
#include <stdlib.h>

#define MAX 120
#define MIN 0
cudaError_t addWithCuda(int *c, const int *a, size_t size);

__global__ void addKernel(int *c, const int *a) {
    int i = threadIdx.x;
    extern __shared__ int smem[];
    smem[i] = a[i];
    __syncthreads();
    if (i == 0)  // 0号线程做平方和
            {
        c[0] = 0;
        for (int d = 0; d < 5; d++) {
            c[0] += smem[d] * smem[d];
        }
    }
    if (i == 1)  //1号线程做累加
            {
        c[1] = 0;
        for (int d = 0; d < 5; d++) {
            c[1] += smem[d];
        }
    }
    if (i == 2)  //2号线程做累乘
            {
        c[2] = 1;
        for (int d = 0; d < 5; d++) {
            c[2] = smem[d];
        }

    }

    if (i == 3)  //3号线程做异或
                {
            c[3] = 0;
            for (int d = 0; d < 5; d++) {
                c[3] ^= smem[d];
            }

        }
}

int main() {
    const int arraySize = 5;
    srand((unsigned) time(NULL));
    const int a[arraySize] = { rand() % (MAX + 1 - MIN) + MIN, rand()
            % (MAX + 1 - MIN) + MIN, rand() % (MAX + 1 - MIN) + MIN, rand()
            % (MAX + 1 - MIN) + MIN, rand() % (MAX + 1 - MIN) + MIN };
    int c[arraySize] = { 0 };
    // Add vectors in parallel.
    cudaError_t cudaStatus = addWithCuda(c, a, arraySize);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "addWithCuda failed!");
        return 1;
    }
    printf(
            "	%d+%d+%d+%d+%d = %d
	%d^2+%d^2+%d^2+%d^2+%d^2 = %d
	%d*%d*%d*%d*%d = %d
	%d^%d^%d^%d^%d = %d




",
            a[0], a[1], a[2], a[3], a[4], c[1], a[0], a[1], a[2], a[3], a[4],
            c[0], a[0], a[1], a[2], a[3], a[4], c[2],a[0], a[1], a[2], a[3], a[4], c[3]);
    // cudaThreadExit must be called before exiting in order for profiling and
    // tracing tools such as Nsight and Visual Profiler to show complete traces.
    cudaStatus = cudaThreadExit();
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaThreadExit failed!");
        return 1;
    }
    return 0;
}

// Helper function for using CUDA to add vectors in parallel.
cudaError_t addWithCuda(int *c, const int *a, size_t size) {
    int *dev_a = 0;
    int *dev_c = 0;
    cudaError_t cudaStatus;

    // Choose which GPU to run on, change this on a multi-GPU system.
    cudaStatus = cudaSetDevice(0);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr,
                "cudaSetDevice failed!  Do you have a CUDA-capable GPU installed?");
        goto Error;
    }

    // Allocate GPU buffers for three vectors (two input, one output)    .
    cudaStatus = cudaMalloc((void**) &dev_c, size * sizeof(int));
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMalloc failed!");
        goto Error;
    }

    cudaStatus = cudaMalloc((void**) &dev_a, size * sizeof(int));
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMalloc failed!");
        goto Error;
    }
    // Copy input vectors from host memory to GPU buffers.
    cudaStatus = cudaMemcpy(dev_a, a, size * sizeof(int),
            cudaMemcpyHostToDevice);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMemcpy failed!");
        goto Error;
    }
    // Launch a kernel on the GPU with one thread for each element.
    addKernel<<<1, size, size * sizeof(int), 0>>>(dev_c, dev_a);

    // cudaThreadSynchronize waits for the kernel to finish, and returns
    // any errors encountered during the launch.
    cudaStatus = cudaThreadSynchronize();
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr,
                "cudaThreadSynchronize returned error code %d after launching addKernel!
",
                cudaStatus);
        goto Error;
    }

    // Copy output vector from GPU buffer to host memory.
    cudaStatus = cudaMemcpy(c, dev_c, size * sizeof(int),
            cudaMemcpyDeviceToHost);
    if (cudaStatus != cudaSuccess) {
        fprintf(stderr, "cudaMemcpy failed!");
        goto Error;
    }

    Error: cudaFree(dev_c);
    cudaFree(dev_a);
    return cudaStatus;
}

22+103+61+63+17 = 266
22^2+103^2+61^2+63^2+17^2 = 19072
22*103*61*63*17 = 17
22^103^61^63^17 = 98