From b400820185db543ff98fd823a15958444e12fee6 Mon Sep 17 00:00:00 2001 From: Tor Aamodt Date: Sun, 8 Aug 2010 21:15:00 -0800 Subject: ** fully decouple kernel compilation from kernel launch for opencl changes required to do this (this enables oclReduction SDK 3.1 to pass and is a prerequisite to Fermi style concurrent kernel launching) - separate kernel launch for cuda and opencl (and init_grid functions) - add pgm_info structure for holding kernel function_info pointer in _cl_program object after ptx compilation ** added support for clEnqueueNDRangeKernel with local_work_size==NULL which is part of OpenCL spec and used in oclSortingNetworks (however, this is a braindead implementation that only handles the case where global_work_size[0] is smaller than the max number of threads per shader; moreover oclSortingNetworks is still not working but for what looks like another reason) ** refactoring / cleanup - g_global_symbol_table made static to ptx_parser.cc - remove g_kernel_name_to_symtab_lookup (not really being used) - moving various function prototypes into headers (e.g., if foo() defined in bar.cc, then place prototype in bar.h) - adding icnt_reg_options() [git-p4: depot-paths = "//depot/gpgpu_sim_research/fermi/distribution/": change = 7168] --- src/gpgpu-sim/icnt_wrapper.h | 10 ++-------- 1 file changed, 2 insertions(+), 8 deletions(-) (limited to 'src/gpgpu-sim/icnt_wrapper.h') diff --git a/src/gpgpu-sim/icnt_wrapper.h b/src/gpgpu-sim/icnt_wrapper.h index 998a4f3..0961fc4 100644 --- a/src/gpgpu-sim/icnt_wrapper.h +++ b/src/gpgpu-sim/icnt_wrapper.h @@ -75,26 +75,20 @@ typedef void (*icnt_transfer_p)( ); typedef unsigned (*icnt_busy_p)( ); typedef void (*icnt_drain_p)( ); - extern icnt_has_buffer_p icnt_has_buffer; extern icnt_push_p icnt_push; extern icnt_pop_p icnt_pop; extern icnt_transfer_p icnt_transfer; extern icnt_busy_p icnt_busy; extern icnt_drain_p icnt_drain; +extern int g_network_mode; -/* -// definition of valid gpu network mode. -extern enum { - INTERSIM = 1, - N_NETWORK_MODE -} gpu_network_mode; -*/ enum network_mode { INTERSIM = 1, N_NETWORK_MODE }; void icnt_init( unsigned int n_shader, unsigned int n_mem ); +void icnt_reg_options( class OptionParser * opp ); #endif -- cgit v1.3