Heth at 10:05, 16 May 2010

2010-05-16T10:05:38Z

Heth at 10:04, 16 May 2010

2010-05-16T10:04:07Z

Heth: CUDA notes 1 moved to CUDA example 1

2010-05-16T09:53:51Z

Heth: New page: // includes, system #include #include // Simple utility function to check for CUDA runtime errors void checkCUDAError(const char* msg); // Part3: imp...

2010-05-16T09:53:37Z

New page: <source lang=C> // includes, system #include <stdio.h> #include <assert.h> // Simple utility function to check for CUDA runtime errors void checkCUDAError(const char* msg); // Part3: imp...

New page

<source lang=C>
// includes, system
#include <stdio.h>
#include <assert.h>

// Simple utility function to check for CUDA runtime errors
void checkCUDAError(const char* msg);

// Part3: implement the kernel
__global__ void reverseArrayBlock(int *d_out, int *d_in)
{
int inOffset = blockDim.x * blockIdx.x;
int outOffset = blockDim.x * (gridDim.x - 1 - blockIdx.x);
int in = inOffset + threadIdx.x;
int out = outOffset + (blockDim.x - 1 - threadIdx.x);
d_out[out] = d_in[in];
}
/////////////////////////////////////////////////////////////////////
// Program main
/////////////////////////////////////////////////////////////////////
int main( int argc, char** argv)
{
// pointer for host memory and size
int *h_a;
int dimA = 256 * 1024; // 256K elements (1MB total)

// pointer for device memory
int *d_b, *d_a;

// define grid and block size
int numThreadsPerBlock = 256;

// Part 1: compute number of blocks needed based on
// array size and desired block size
int numBlocks = dimA / numThreadsPerBlock;

// allocate host and device memory
size_t memSize = numBlocks * numThreadsPerBlock * sizeof(int);
h_a = (int *) malloc(memSize);
cudaMalloc( (void **) &d_a, memSize );
cudaMalloc( (void **) &d_b, memSize );

// Initialize input array on host
for (int i = 0; i < dimA; ++i)
{
h_a[i] = i;
}

// Copy host array to device array
cudaMemcpy( d_a, h_a, memSize, cudaMemcpyHostToDevice );

// launch kernel
dim3 dimGrid(numBlocks);
dim3 dimBlock(numThreadsPerBlock);
reverseArrayBlock<<< dimGrid, dimBlock >>>( d_b, d_a );

// block until the device has completed
cudaThreadSynchronize();

// check if kernel execution generated an error
// Check for any CUDA errors
checkCUDAError("kernel invocation");

// device to host copy
cudaMemcpy( h_a, d_b, memSize, cudaMemcpyDeviceToHost );

// Check for any CUDA errors
checkCUDAError("memcpy");

// verify the data returned to the host is correct
for (int i = 0; i < dimA; i++)
{
assert(h_a[i] == dimA - 1 - i );
}

// free device memory
cudaFree(d_a);
cudaFree(d_b);

// free host memory
free(h_a);

// If the program makes it this far, then the results are
// correct and there are no run-time errors. Good work!
printf("Correct!\n");

return 0;
}
void checkCUDAError(const char *msg)
{
cudaError_t err = cudaGetLastError();
if( cudaSuccess != err)
{
fprintf(stderr, "Cuda error: %s: %s.\n", msg,
cudaGetErrorString( err) );
exit(EXIT_FAILURE);
}
}
</source>

← Older revision		Revision as of 10:04, 16 May 2010
Line 98:		Line 98:
	}		}
	</source>		</source>
		+	[[Category:CUDA]]

@@ Line 1: / Line 1: @@
-<source lang=C>
+<source lang=CUDA>
 // includes, system
 #include <stdio.h>

← Older revision	Revision as of 09:53, 16 May 2010
(No difference)

CUDA example 1 - Revision history

Heth at 10:05, 16 May 2010

Heth at 10:04, 16 May 2010

Heth: CUDA notes 1 moved to CUDA example 1

Heth: New page: // includes, system #include #include // Simple utility function to check for CUDA runtime errors void checkCUDAError(const char* msg); // Part3: imp...