From 6a39e5c5964f23a97dafaa6a66f2a9d9c37bbfdd Mon Sep 17 00:00:00 2001 From: Jonathan Date: Fri, 18 May 2018 11:52:31 -0700 Subject: implemented unbind, currently only affects cudaArray map --- 0517_13-24cudaReserved.log | 0 517-1418cudaReserved.log | 0 libcuda/cuda_runtime_api.cc | 11 +++++++++++ src/abstract_hardware_model.h | 1 + src/cuda-sim/cuda-sim.cc | 5 +++++ 5 files changed, 17 insertions(+) delete mode 100644 0517_13-24cudaReserved.log delete mode 100644 517-1418cudaReserved.log diff --git a/0517_13-24cudaReserved.log b/0517_13-24cudaReserved.log deleted file mode 100644 index e69de29..0000000 diff --git a/517-1418cudaReserved.log b/517-1418cudaReserved.log deleted file mode 100644 index e69de29..0000000 diff --git a/libcuda/cuda_runtime_api.cc b/libcuda/cuda_runtime_api.cc index e3c2542..71926f8 100644 --- a/libcuda/cuda_runtime_api.cc +++ b/libcuda/cuda_runtime_api.cc @@ -991,6 +991,14 @@ __host__ cudaError_t CUDARTAPI cudaBindTextureToArray(const struct textureRefere __host__ cudaError_t CUDARTAPI cudaUnbindTexture(const struct textureReference *texref) { + CUctx_st *context = GPGPUSim_Context(); + gpgpu_t *gpu = context->get_device()->get_gpgpu(); + printf("GPGPU-Sim PTX: in cudaUnbindTexture: sizeof(struct textureReference) = %zu\n", sizeof(struct textureReference)); + printf("GPGPU-Sim PTX: Name corresponding to textureReference: %s\n", gpu->gpgpu_ptx_sim_findNamefromTexture(texref)); + const struct cudaArray* array = gpu->get_texarray(texref); + printf("GPGPU-Sim PTX: texref = %p, array = %p\n", texref, array); + + gpu->gpgpu_ptx_sim_unbindTexture(texref); return g_last_cudaError = cudaSuccess; } @@ -2073,10 +2081,12 @@ void __cudaUnregisterFatBinary(void **fatCubinHandle) cudaError_t cudaDeviceReset ( void ) { // Should reset the simulated GPU + // TODO: Implement return g_last_cudaError = cudaSuccess; } cudaError_t CUDARTAPI cudaDeviceSynchronize(void){ // I don't know what this should do + // TODO: Implement return g_last_cudaError = cudaSuccess; } @@ -2178,6 +2188,7 @@ typedef unsigned long GLuint; cudaError_t cudaGLRegisterBufferObject(GLuint bufferObj) { printf("GPGPU-Sim PTX: Execution warning: ignoring call to \"%s\"\n", __my_func__ ); + // TODO: Implement return g_last_cudaError = cudaSuccess; } diff --git a/src/abstract_hardware_model.h b/src/abstract_hardware_model.h index 412c0a8..3ef450e 100644 --- a/src/abstract_hardware_model.h +++ b/src/abstract_hardware_model.h @@ -544,6 +544,7 @@ public: class memory_space *get_surf_memory() { return m_surf_mem; } void gpgpu_ptx_sim_bindTextureToArray(const struct textureReference* texref, const struct cudaArray* array); + void gpgpu_ptx_sim_unbindTexture(const struct textureReference* texref); void gpgpu_ptx_sim_bindNameToTexture(const char* name, const struct textureReference* texref, int dim, int readmode, int ext); const char* gpgpu_ptx_sim_findNamefromTexture(const struct textureReference* texref); diff --git a/src/cuda-sim/cuda-sim.cc b/src/cuda-sim/cuda-sim.cc index 946043a..ef16f43 100644 --- a/src/cuda-sim/cuda-sim.cc +++ b/src/cuda-sim/cuda-sim.cc @@ -183,6 +183,11 @@ void gpgpu_t::gpgpu_ptx_sim_bindTextureToArray(const struct textureReference* te m_TextureRefToTexureInfo[texref] = texInfo; } +void gpgpu_t::gpgpu_ptx_sim_unbindTexture(const struct textureReference* texref) +{ + m_TextureRefToCudaArray.erase(texref); +} + unsigned g_assemble_code_next_pc=0; std::map g_pc_to_finfo; std::vector function_info::s_g_pc_to_insn; -- cgit v1.3