diff options
| author | Andrew <saintruler@gmail.com> | 2019-05-05 22:53:12 +0400 |
|---|---|---|
| committer | Andrew <saintruler@gmail.com> | 2019-05-05 22:53:12 +0400 |
| commit | 3a94328b3cff255ac64a208ac98919cc3e530aa0 (patch) | |
| tree | 377ab395a13e95b5c12a442645c2eaa4b2a644f8 /Canvas/gpu/CalculateCanvas.cu | |
| parent | 15ccb946f43283bfc91d80ddf012d9ae2a0333ed (diff) | |
Made changes to CMakeLists.txt so CUDA version of program builds when CUDA is installed.
Replaced C-style approach to generating texture to OOP-style with Canvas.h
Diffstat (limited to 'Canvas/gpu/CalculateCanvas.cu')
| -rw-r--r-- | Canvas/gpu/CalculateCanvas.cu | 52 |
1 files changed, 52 insertions, 0 deletions
diff --git a/Canvas/gpu/CalculateCanvas.cu b/Canvas/gpu/CalculateCanvas.cu new file mode 100644 index 0000000..3a73067 --- /dev/null +++ b/Canvas/gpu/CalculateCanvas.cu @@ -0,0 +1,52 @@ +#include "CalculateCanvas.h" + +__global__ +void calculate(unsigned char* canvas, int width, Point* points, int nPoints) +{ + int canvasX = blockIdx.x * blockDim.x + threadIdx.x; + int canvasY = blockIdx.y * blockDim.y + threadIdx.y; + + int pathX, pathY; + int length_squared; + double f = 0; + + for (int i = 0; i < nPoints; i++) + { + pathX = points[i].x - canvasX; + pathY = points[i].y - canvasY; + + length_squared = pathX * pathX + pathY * pathY; + f += (1e5 * 20) / (float) length_squared; + } + + int l = (int) f; + if (l > 255) l = 255; + + canvas[canvasY * width * 3 + canvasX * 3 + 0] = l; + canvas[canvasY * width * 3 + canvasX * 3 + 1] = 0; + canvas[canvasY * width * 3 + canvasX * 3 + 2] = 0; +} + +void generate_canvas(int width, int height, unsigned char* canvas, Point* points, int nPoints) +{ + const int canvasSize = width * height * 3 * sizeof(unsigned char); + const int pointsSize = nPoints * sizeof(Point); + + unsigned char* gpuCanvas; + Point* gpuPoints; + + cudaMalloc((void**) &gpuCanvas, canvasSize); + cudaMalloc((void**) &gpuPoints, pointsSize); + + cudaMemcpy(gpuCanvas, canvas, canvasSize, cudaMemcpyHostToDevice); + cudaMemcpy(gpuPoints, points, pointsSize, cudaMemcpyHostToDevice); + + dim3 threadsPerBlock(16, 16); + dim3 numBlocks(width / threadsPerBlock.x, height / threadsPerBlock.y); + calculate<<<numBlocks, threadsPerBlock>>>(gpuCanvas, width, gpuPoints, nPoints); + + cudaMemcpy(canvas, gpuCanvas, canvasSize, cudaMemcpyDeviceToHost); + + cudaFree(gpuCanvas); + cudaFree(gpuPoints); +} |