dx = blockIdx.x * blockDim.x + threadIdx.x;
// Копіюємо з глобальної в локальну блоками
data [tid] = (idx
// Чекаємо поки кожен потік в блоці скопіює дані
__syncthreads ();
// Підсумовування в паралелі (int s = blockDim.x/2; s> 0; s>> = 1)
{(tid
__syncthreads ();
}
// Перший потік в блоці записує результат підсумовування
if (tid == 0) _Dst [blockIdx.x] = data [0];
}
extern "C" void reductionGPU (int * d_Dst, int * d_Src)
{<<>> (d_Dst, d_Src); ("convolutionRowsKernel () execution failed n");
}
# include
# include
# include
# include
# include
# include
# include "common.h"
////////////////////////////////////////////////////////////////////////////////
// Main program
////////////////////////////////////////////////////////////////////////////////main (int argc, char ** argv)
{
// Use command-line specified CUDA device, otherwise use device with highest Gflops/s (cutCheckCmdLineFlag (argc, (const char **) argv, "device")) (argc, argv); (cutGetMaxGflopsDeviceId () );
int * d_Input, * d_Output; * h_Input = new int [VECTOR_SIZE]; * h_Output = new int [BLOCKS];
(int i = 0; i
{_Input [i] = 1;
}
std :: cout <<"Size of data:" <
std :: cout <<"Reading back GPU results ... n"; (cudaMemcpy (h_Output, d_Output, BLOCKS * sizeof (int), cudaMemcpyDeviceToHost)); :: cout <<" ; Results: n "; :: copy (h_Output, h_Output + BLOCKS, std :: ostream_iterator (std :: cout,", ")); (cudaFree (d_Input)); (cudaFree (d_Output) ); [] h_Input; [] h_Output; ();
cutilExit (argc, argv);}
Технічний висновок программиof data: 8192: 16and initializing CUDA arrays ... GPU reduction ... back GPU results ...:
, 512, 512, 512, 512, 512, 512, 512, 512, 512, 512, 512, 512, 512, 512, 512ENTER to exit ...
Код програми підсумовування вектора CPU
# include
# include
# include
# include
# include show (const std :: ...