CUDA 做排列组合

#include <fstream>#include <stdio.h>#include <stdlib.h>#include <cuda_runtime.h>#include <iostream>using namespace std;#define ROWS 10#define COLS 6#define CHECK(res) if(res!=cudaSuccess){exit(-1);}__global__ void Kerneltest(double **dp_out_params, double * d_out_Iqsd1, double *d_MTFac, double *d_errIqsd_MTFac, unsigned int loop){    unsigned int row =  blockDim.x*blockIdx.x + threadIdx.x   ;    unsigned int col = threadIdx.y;double min_q = 0;double max_q = 0.027;int stepNums_q = 10;double stepLen_q = (max_q-min_q)/stepNums_q;double min_mua = 0;double max_mua = 0.007;int stepNums_mua = 100;double stepLen_mua = (max_mua-min_mua)/stepNums_mua;double min_musp = 0;double max_musp = 1;int stepNums_musp = 100;double stepLen_musp = (max_musp-min_musp)/stepNums_musp;double min_alpha = 0;double max_alpha = 1;int stepNums_alpha = 100;double stepLen_alpha = (max_alpha-min_alpha)/stepNums_alpha;double min_n = 0;double max_n = 1;int stepNums_n = 100;double stepLen_n = (max_n-min_n)/stepNums_n;double min_rough = 0;double max_rough = 1;int stepNums_rough = 100;double stepLen_rough = (max_rough-min_rough)/stepNums_rough;if (row < ROWS && col < COLS){dp_out_params[row][col] = 0; __syncthreads();if( col== 5 ){int weight_q = (row+ROWS*loop) % stepNums_q;dp_out_params[row][col] = min_q + weight_q * stepLen_q;} __syncthreads();if( col== 4 ){int weight_mua =( (row+ROWS*loop) / stepNums_q ) % stepNums_mua;dp_out_params[row][col] = min_mua + weight_mua * stepLen_mua;} __syncthreads();if( col== 3 ){int weight_musp =( (row+ROWS*loop) / stepNums_q / stepNums_mua ) % stepNums_musp;dp_out_params[row][col] = min_musp + weight_musp * stepLen_musp; } __syncthreads();if( col== 2 ){int weight_alpha =( (row+ROWS*loop) / stepNums_q / stepNums_mua /stepNums_musp ) % stepNums_alpha;dp_out_params[row][col] = min_alpha + weight_alpha * stepLen_alpha; } __syncthreads();if( col== 1 ){int weight_n =( (row+ROWS*loop) / stepNums_q / stepNums_mua /stepNums_musp / stepNums_alpha ) % stepNums_n;dp_out_params[row][col] = min_n + weight_n * stepLen_n; } __syncthreads();if( col== 0 ){int weight_rough =( (row+ROWS*loop) / stepNums_q / stepNums_mua /stepNums_musp / stepNums_alpha / stepNums_n ) % stepNums_rough;dp_out_params[row][col] = min_rough + weight_rough * stepLen_rough; } __syncthreads();}///////////////////////////////////////////////////////////////////////////////////d_out_Iqsd1[row] = threadIdx.x;__syncthreads();///////////////////////////////////////////////////////////////////////////////////    unsigned int row1 =  blockDim.x*blockIdx.x + threadIdx.x   ;d_errIqsd_MTFac[row1] = d_out_Iqsd1[row] - d_MTFac[row];__syncthreads();}//////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////////int main(int argc, char **argv){cudaError_t res;////////////////////////////////////////int MTFlen = 4000;double *h_MTFac = NULL;h_MTFac = (double*)malloc(MTFlen*sizeof(double));for(int i=0; i<100; i++){h_MTFac[i] = 3.2;}double *d_MTFac = NULL;    res = cudaMalloc((void**)(&d_MTFac), MTFlen*sizeof(double));CHECK(res)    res = cudaMemcpy((void*)(d_MTFac), (void*)(h_MTFac), MTFlen*sizeof(double*), cudaMemcpyHostToDevice);CHECK(res)double *h_errIqsd_MTFac = NULL;    h_errIqsd_MTFac = (double*)malloc(MTFlen*sizeof(double));double *d_errIqsd_MTFac = NULL;    res = cudaMalloc((void**)(&d_errIqsd_MTFac), MTFlen*sizeof(double));CHECK(res)    res = cudaMemcpy((void*)(d_errIqsd_MTFac), (void*)(h_errIqsd_MTFac), MTFlen*sizeof(double*), cudaMemcpyHostToDevice);CHECK(res)////////////////////////////////////////double *d_out_params = NULL;    res = cudaMalloc((void**)(&d_out_params), ROWS*COLS*sizeof(double));CHECK(res)double **dp_out_params = NULL;res = cudaMalloc((void**)(&dp_out_params), ROWS*sizeof(double*));CHECK(res)double **hp_out_params = NULL;hp_out_params = (double**)malloc(ROWS*sizeof(double*));double *h_out_params = NULL;h_out_params = (double*)malloc(ROWS*COLS*sizeof(double));    for (int r = 0; r < ROWS; r++)    {    hp_out_params[r] = d_out_params + r*COLS;    }////////////////////////////////////////    double *h_out_Iqsd1 = NULL;    h_out_Iqsd1 = (double*)malloc(ROWS*sizeof(double));double *d_out_Iqsd1 = NULL;res = cudaMalloc((void **) &d_out_Iqsd1, ROWS*sizeof(double));CHECK(res)////////////////////////////////////////    res = cudaMemcpy((void*)(dp_out_params), (void*)(hp_out_params), ROWS*sizeof(double*), cudaMemcpyHostToDevice);CHECK(res)    dim3 dimBlock( 15,   6,  1);    dim3 dimGrid(  1,    1,  1);    for(unsigned int loop=0; loop<1; loop++)     {    Kerneltest<<<dimGrid, dimBlock>>>(dp_out_params, d_out_Iqsd1, d_MTFac, d_errIqsd_MTFac, loop);    cout<<"loop:  "<<loop<<endl;     }    res = cudaMemcpy((void*)(h_out_params), (void*)(d_out_params), ROWS*COLS*sizeof(double*), cudaMemcpyDeviceToHost);CHECK(res)    res = cudaMemcpy((void*)(h_out_Iqsd1), (void*)(d_out_Iqsd1), ROWS*sizeof(double*), cudaMemcpyDeviceToHost);CHECK(res)    res = cudaMemcpy((void*)(h_MTFac), (void*)(d_MTFac), MTFlen*sizeof(double*), cudaMemcpyDeviceToHost);CHECK(res)    res = cudaMemcpy((void*)(h_errIqsd_MTFac), (void*)(d_errIqsd_MTFac), MTFlen*sizeof(double*), cudaMemcpyDeviceToHost);CHECK(res)////////////////////////////////////////    ofstream f1("/home/zlf/Documents/cuda.txt");    int zz = 0;cout<<endl<<"h_out_params: "<<endl;    for (int r = 0; r < ROWS; r++)    {        for (int c = 0; c < COLS; c++)        {            printf("%f ", h_out_params[r*COLS+c]);            f1 << h_out_params[r*COLS+c]<<"    ";        }        zz = zz + 1;        cout<<"    行数： "<<zz;        printf("\n");        f1<< "    行数： "<<zz<< "\n";    }f1.close();cout<<endl<<"h_out_Iqsd1: "<<endl;    for (int r = 0; r < ROWS; r++)    {    cout<<h_out_Iqsd1[r]<<"    ";    if ((r%10)==9)    {    cout<<endl;    }    }    /////////////////////////    cout<<"h_errIqsd_MTFac[i]: "<<endl;for(int i=0; i<MTFlen; i++){cout<<h_errIqsd_MTFac[i]<<"  ";}    /////////////////////////cout<<zz<<endl;    cudaFree((void*)d_out_params);    cudaFree((void*)dp_out_params);    cudaFree((void*)d_out_Iqsd1);    cudaFree((void*)d_MTFac);    cudaFree((void*)d_errIqsd_MTFac);    free(h_out_params);    free(hp_out_params);    free(h_out_Iqsd1);    free(h_MTFac);    free(h_errIqsd_MTFac);    getchar();    return 0;}