From 61aaff767ef7009d04510dec9ceabf49c2cc087d Mon Sep 17 00:00:00 2001 From: NagithaAbeywickrema Date: Tue, 25 Apr 2023 12:37:17 +0530 Subject: [PATCH] Add documentation to internal structures --- src/cuda.c | 14 +++++--- src/hip.cpp | 16 ++++++--- src/ispc.c | 26 ++++++++++----- src/jit.c | 7 ++-- src/log.c | 9 +++-- src/nomp-impl.h | 89 ++++++++++++++++++++++++++++++------------------- src/opencl.c | 16 ++++++--- src/sycl.cpp | 17 +++++++--- 8 files changed, 127 insertions(+), 67 deletions(-) diff --git a/src/cuda.c b/src/cuda.c index ccbdf3a4..dc4cafad 100644 --- a/src/cuda.c +++ b/src/cuda.c @@ -35,14 +35,20 @@ static const char *ERR_STR_CUDA_FAILURE = "Cuda %s failed: %s."; } \ } +/** + * @brief Struct to store device ID and device properties for a CUDA backend. + */ struct cuda_backend { - int device_id; - struct cudaDeviceProp prop; + int device_id; /**< Device ID of the CUDA backend */ + struct cudaDeviceProp prop; /**< Device properties of the CUDA backend */ }; +/** + * @brief Struct to store CUDA module and kernel function + */ struct cuda_prog { - CUmodule module; - CUfunction kernel; + CUmodule module; /**< CUDA module containing the kernel function */ + CUfunction kernel; /**< CUDA kernel function */ }; static int cuda_update(struct backend *bnd, struct mem *m, const int op) { diff --git a/src/hip.cpp b/src/hip.cpp index 1157b16b..d09eac8e 100644 --- a/src/hip.cpp +++ b/src/hip.cpp @@ -27,15 +27,21 @@ const char *ERR_STR_HIP_FAILURE = "HIP %s failed: %s."; +/** + * @brief Struct to store HIP backend specific data. + */ struct hip_backend { - int device_id; - struct hipDeviceProp_t prop; - hipCtx_t ctx; + int device_id; /**< Device ID of the HIP backend */ + struct hipDeviceProp_t prop; /**< Device properties of the HIP backend */ + hipCtx_t ctx; /**< Context of the HIP backend (Used in Nvidia devices only) */ }; +/** + * @brief Struct to store the HIP module and kernel function. + */ struct hip_prog { - hipModule_t module; - hipFunction_t kernel; + hipModule_t module; /**< HIP module of the compiled program */ + hipFunction_t kernel; /**< Kernel function of the compiled program */ }; static int hip_update(struct backend *bnd, struct mem *m, const int op) { diff --git a/src/ispc.c b/src/ispc.c index 440c9a1a..251bb978 100644 --- a/src/ispc.c +++ b/src/ispc.c @@ -5,18 +5,28 @@ static const char *ERR_STR_ISPC_FAILURE = "ISPC %s failed with error message: %s."; +/** + * @brief Struct to store information about the ISPC backend. + */ struct ispc_backend { - char *ispc_cc, *cc, *ispc_flags, *cc_flags; - ISPCRTDevice device; - ISPCRTDeviceType device_type; - ISPCRTTaskQueue queue; - ISPCRTNewMemoryViewFlags flags; + char *ispc_cc; /**< Path to the ISPC compiler */ + char *cc; /**< Path to the C/C++ compiler */ + char *ispc_flags; /**< ISPC compiler flags */ + char *cc_flags; /**< C/C++ compiler flags */ + ISPCRTDevice device; /**< Handle to the selected device */ + ISPCRTDeviceType device_type; /**< Type of device */ + ISPCRTTaskQueue queue; /**< Handle to the task queue */ + ISPCRTNewMemoryViewFlags + flags; /**< Flags to configure the creation of new memory views */ }; +/** + * @brief Struct to store ISPC program information. + */ struct ispc_prog { - int ispc_id; - ISPCRTModule module; - ISPCRTKernel kernel; + int ispc_id; /**< ID of the ISPC program */ + ISPCRTModule module; /**< Module of the ISPC program */ + ISPCRTKernel kernel; /**< Kernel of the ISPC program */ }; static ISPCRTError rt_error = ISPCRT_NO_ERROR; diff --git a/src/jit.c b/src/jit.c index 0a8f163f..656699a0 100644 --- a/src/jit.c +++ b/src/jit.c @@ -102,9 +102,12 @@ static int compile_aux(const char *cc, const char *cflags, const char *src, return failed; } +/** + * @brief Struct to store information about a dynamic loaded function. + */ struct function { - void *dlh; - void (*dlf)(void **); + void *dlh; /**< Handle to the dynamic library that contains the function */ + void (*dlf)(void **); /**< Pointer to the function */ }; static struct function **funcs = NULL; diff --git a/src/log.c b/src/log.c index 5a15d653..06b29336 100644 --- a/src/log.c +++ b/src/log.c @@ -10,10 +10,13 @@ const char *ERR_STR_USER_MAP_PTR_IS_INVALID = const char *ERR_STR_USER_DEVICE_IS_INVALID = "Device id %d passed into libnomp is not valid."; +/** + * @brief Struct to store information about a log entry. + */ struct log { - char *description; - int logno; - nomp_log_type type; + char *description; /**< Description of the log entry */ + int logno; /**< Log number associated with the entry */ + nomp_log_type type; /**< Type of log */ }; static struct log *logs = NULL; diff --git a/src/nomp-impl.h b/src/nomp-impl.h index d4e0e8de..cec06260 100644 --- a/src/nomp-impl.h +++ b/src/nomp-impl.h @@ -31,51 +31,70 @@ #include "nomp.h" +/** + * @brief Represents a memory block that can be used for data storage and + * transfer between a host and a device. + */ struct mem { - size_t idx0, idx1, usize; - void *hptr, *bptr; - size_t bsize; + size_t idx0; /**< Starting index of a memory block */ + size_t idx1; /**< Ending index of a memory block */ + size_t usize; /**< Size (in bytes) of each element in the memory block */ + void *hptr; /**< Pointer to the host memory */ + void *bptr; /**< Pointer to the device memory */ + size_t bsize; /**< Size (in bytes) of allocated memory on device */ }; +/** + * @brief Represents a kernel argument. + */ struct arg { - char name[MAX_ARG_NAME_SIZE]; - size_t size; - unsigned type; - void *ptr; + char name[MAX_ARG_NAME_SIZE]; /**< The name of the argument */ + size_t size; /**< The size of the argument data */ + unsigned type; /**< The type of the argument */ + void *ptr; /**< A pointer to the argument data */ }; +/** + * @brief Struct to store meta information about kernel arguments. + */ struct prog { - // Number of arguments of the kernel and meta info about - // arguments. - unsigned nargs; - struct arg *args; - // Dimension of kernel launch parameters, their pymbolic - // expressions, and evaluated value of each dimension. - unsigned ndim; - PyObject *py_global, *py_local; - size_t global[3], local[3]; - // Map of variable names and their values use to evaluate - // the kernel launch parameters. - PyObject *py_dict; - // Pointer to keep track of backend specific data. - void *bptr; + unsigned nargs; /**< Number of kernel arguments */ + struct arg *args; /**< Pointer to an array of kernel arguments */ + unsigned ndim; /**< Dimension of kernel launch parameters */ + PyObject *py_global; /**< Pymbolic expressions for global dimensions */ + PyObject *py_local; /**< Pymbolic expressions for local dimensions */ + size_t global[3]; /**< Sizes of each global dimension */ + size_t local[3]; /**< Sizes of each local dimension */ + PyObject *py_dict; /**< Map of variable names to their values used to evaluate + the kernel launch parameters */ + void *bptr; /**< Pointer to backend specific data */ }; +/** + * @brief Struct to store user configurations and pointers to backend functions. + */ struct backend { - // User configurations of the backend. - int platform_id, device_id, verbose; - char *backend, *install_dir; - // Python function object which will be called to perform annotations. - PyObject *py_annotate; - // Pointers to backend functions used for backend dispatch. - int (*update)(struct backend *, struct mem *, const int); - int (*knl_build)(struct backend *, struct prog *, const char *, const char *); - int (*knl_run)(struct backend *, struct prog *); - int (*knl_free)(struct prog *); - int (*sync)(struct backend *); - int (*finalize)(struct backend *); - // Pointer to keep track of backend specific data. - void *bptr; + int platform_id; /**< Platform ID of the backend */ + int device_id; /**< Device ID of the backend */ + int verbose; /**< Verbosity level */ + char *backend; /**< Name of the backend */ + char *install_dir; /**< Nomp installation directory */ + PyObject *py_annotate; /**< Python function object which will be called to + perform annotations */ + int (*update)(struct backend *, struct mem *, + const int); /**< Pointer to backend memory update function */ + int (*knl_build)( + struct backend *, struct prog *, const char *, + const char *); /**< Pointer to backend kernel build function */ + int (*knl_run)(struct backend *, + struct prog *); /**< Pointer to backend kernel run function */ + int (*knl_free)( + struct prog *); /**< Pointer to backend kernel free function */ + int (*sync)( + struct backend *); /**< Pointer to backend synchronization function */ + int (*finalize)( + struct backend *); /**< Pointer to backend finalizing function */ + void *bptr; /**< Pointer to backend specific data */ }; #ifdef __cplusplus diff --git a/src/opencl.c b/src/opencl.c index 112bfb3c..0ed15c29 100644 --- a/src/opencl.c +++ b/src/opencl.c @@ -18,15 +18,21 @@ static const char *ERR_STR_OPENCL_FAILURE = "%s failed with error code: %d."; } \ } +/** + * @brief Struct to store OpenCL backend-specific data. + */ struct opencl_backend { - cl_device_id device_id; - cl_command_queue queue; - cl_context ctx; + cl_device_id device_id; /**< OpenCL device ID */ + cl_command_queue queue; /**< OpenCL command queue */ + cl_context ctx; /**< OpenCL context */ }; +/** + * @brief Struct to store an OpenCL program and kernel. + */ struct opencl_prog { - cl_program prg; - cl_kernel knl; + cl_program prg; /**< OpenCL program */ + cl_kernel knl; /**< OpenCL kernel */ }; static int opencl_update(struct backend *bnd, struct mem *m, const int op) { diff --git a/src/sycl.cpp b/src/sycl.cpp index f1972fbc..d58e86bd 100644 --- a/src/sycl.cpp +++ b/src/sycl.cpp @@ -12,15 +12,22 @@ static const char *ERR_STR_SYCL_FAILURE = "SYCL backend failed with error: %s."; ex.what()); \ } +/** + * @brief Struct to store SYCL backend specific data. + */ struct sycl_backend { - sycl::device device_id; - sycl::queue queue; - sycl::context ctx; - char *compiler, *compiler_flags; + sycl::device device_id; /**< SYCL device ID */ + sycl::queue queue; /**< SYCL queue */ + sycl::context ctx; /**< SYCL context */ + char *compiler; /**< SYCL compiler path */ + char *compiler_flags; /**< SYCL compiler flags */ }; +/** + * @brief Struct to store SYCL program information. + */ struct sycl_prog { - int sycl_id; + int sycl_id; /**< ID of the SYCL program */ }; static int sycl_update(struct backend *bnd, struct mem *m, const int op) {