diff options
| author | Tor Aamodt <[email protected]> | 2010-10-01 08:55:28 -0800 |
|---|---|---|
| committer | Tor Aamodt <[email protected]> | 2010-10-01 08:55:28 -0800 |
| commit | 11b308e7363e937966b035b4891db32b4eece3bf (patch) | |
| tree | 50ca4c9ad6f163ac4acb2bf505e64dfebed66947 /benchmarks/CUDA/template/template_kernel.cu | |
| parent | bb820c116764d7a1b8e071137d32b74e7f34dd2f (diff) | |
integrating recent changes from fermi-test into fermi
(i'll use "fermi" for more disruptive changes to the pipeline model such
as updating the MSHRs and getting rid of the warp tracker, ripping out DWF, etc...)
[git-p4: depot-paths = "//depot/gpgpu_sim_research/fermi/distribution/": change = 7805]
Diffstat (limited to 'benchmarks/CUDA/template/template_kernel.cu')
| -rw-r--r-- | benchmarks/CUDA/template/template_kernel.cu | 66 |
1 files changed, 0 insertions, 66 deletions
diff --git a/benchmarks/CUDA/template/template_kernel.cu b/benchmarks/CUDA/template/template_kernel.cu deleted file mode 100644 index adac694..0000000 --- a/benchmarks/CUDA/template/template_kernel.cu +++ /dev/null @@ -1,66 +0,0 @@ -/* - * Copyright 1993-2010 NVIDIA Corporation. All rights reserved. - * - * NVIDIA Corporation and its licensors retain all intellectual property and - * proprietary rights in and to this software and related documentation. - * Any use, reproduction, disclosure, or distribution of this software - * and related documentation without an express license agreement from - * NVIDIA Corporation is strictly prohibited. - * - * Please refer to the applicable NVIDIA end user license agreement (EULA) - * associated with this source code for terms and conditions that govern - * your use of this NVIDIA software. - * - */ - -/* Template project which demonstrates the basics on how to setup a project - * example application. - * Device code. - */ - -#ifndef _TEMPLATE_KERNEL_H_ -#define _TEMPLATE_KERNEL_H_ - -#include <stdio.h> - -#define SDATA( index) cutilBankChecker(sdata, index) - -//////////////////////////////////////////////////////////////////////////////// -//! Simple test kernel for device functionality -//! @param g_idata input data in global memory -//! @param g_odata output data in global memory -//////////////////////////////////////////////////////////////////////////////// - -__device__ float foo( float bar ) -{ - return bar+1.0f; -} - -__global__ void -testKernel( float* g_idata, float* g_odata) -{ - // shared memory - // the size is determined by the host application - extern __shared__ float sdata[]; - - // access thread id - const unsigned int tid = threadIdx.x; - // access number of threads in this block - const unsigned int num_threads = blockDim.x; - - // read in input data from global memory - // use the bank checker macro to check for bank conflicts during host - // emulation - SDATA(tid) = foo(g_idata[tid])-1.0f; - __syncthreads(); - printf("thread tid=%u reads %f from g_idata[]\n", tid, sdata[tid]); - - // perform some computations - SDATA(tid) = (float) num_threads * SDATA( tid); - __syncthreads(); - - // write data to global memory - g_odata[tid] = SDATA(tid); -} - -#endif // #ifndef _TEMPLATE_KERNEL_H_ |
