diff --git a/dkms.conf b/dkms.conf index f5b504e..d1c7440 100644 --- a/dkms.conf +++ b/dkms.conf @@ -1,5 +1,5 @@ PACKAGE_NAME=aws-neuronx -PACKAGE_VERSION=2.28.0.0 +PACKAGE_VERSION=2.29.0.0 BUILT_MODULE_NAME[0]="neuron" MAKE[0]="make -C ${kernel_source_dir} M=${dkms_tree}/${PACKAGE_NAME}/${PACKAGE_VERSION}/build" CLEAN="make -C ${kernel_source_dir} M=${dkms_tree}/${PACKAGE_NAME}/${PACKAGE_VERSION}/build clean" diff --git a/neuron_arch.c b/neuron_arch.c index f3a6763..04f77ad 100644 --- a/neuron_arch.c +++ b/neuron_arch.c @@ -90,3 +90,20 @@ int narch_get_instance_type_name(char *instance_type_name, size_t instance_type_ return -ENOSYS; #endif } + +enum neuron_platform_type narch_detect_platform_type(const struct neuron_platform_lookup *table) { + char buf[128]; + int i; + + if (narch_get_instance_type_name(buf, sizeof(buf))) { + return NEURON_PLATFORM_TYPE_STD; + } + + for (i = 0; table[i].inst_name != NULL; i++) { + if (strncmp(buf, table[i].inst_name, strlen(table[i].inst_name)) == 0) { + return table[i].platform_type; + } + } + + return NEURON_PLATFORM_TYPE_STD; +} diff --git a/neuron_arch.h b/neuron_arch.h index 831679d..3d4b4cc 100644 --- a/neuron_arch.h +++ b/neuron_arch.h @@ -21,9 +21,15 @@ enum neuron_platform_type { NEURON_PLATFORM_TYPE_STD = 0, NEURON_PLATFORM_TYPE_ULTRASERVER = 1, NEURON_PLATFORM_TYPE_PDS = 2, + NEURON_PLATFORM_TYPE_MAX = 3, NEURON_PLATFORM_TYPE_INVALID, }; +struct neuron_platform_lookup { + const char *inst_name; + enum neuron_platform_type platform_type; +}; + enum neuron_platform_operation_type { NEURON_PLATFORM_OP_TYPE_DEVOPEN = 0, NEURON_PLATFORM_OP_TYPE_EXEC = 1, @@ -83,4 +89,13 @@ bool narch_is_emu(void); */ int narch_get_instance_type_name(char *instance_type_name, size_t instance_type_name_size); +/** + * narch_detect_platform_type() - Detect platform type from instance name. + * + * @table: Array of instance name to platform type mappings. + * + * Return: Matching platform type, or NEURON_PLATFORM_TYPE_STD if no match. + */ +enum neuron_platform_type narch_detect_platform_type(const struct neuron_platform_lookup *table); + #endif diff --git a/neuron_cdev.c b/neuron_cdev.c index dbc12af..e64679e 100644 --- a/neuron_cdev.c +++ b/neuron_cdev.c @@ -20,6 +20,7 @@ #include #include #include +#include #include #include "neuron_ioctl.h" @@ -65,9 +66,12 @@ static struct ncdev devnodes[NEURON_MAX_DEV_NODES]; static int ncdev_mem_chunk_to_mem_handle(struct neuron_device *nd, struct mem_chunk *mc, u64 *mh) { int ret = 0; - if (mc->mc_handle == NMCH_INVALID_HANDLE) { + // Fast path: handle already allocated. A stale INVALID read here is + // harmless — nmch_handle_alloc re-checks under nd->nmch.lock and is + // idempotent per-mc, so a concurrent caller can't leak a second handle. + if (READ_ONCE(mc->mc_handle) == NMCH_INVALID_HANDLE) { ret = nmch_handle_alloc(nd, mc, &mc->mc_handle); - } + } *mh = (u64)mc->mc_handle; return ret; } @@ -84,12 +88,20 @@ static struct mem_chunk *ncdev_mem_handle_to_mem_chunk(struct neuron_device *nd, } -static unsigned long neuron_copy_from_user(const char *const fname, void * to, const void __user * from, unsigned long n) { - const long ret = copy_from_user(to, from, n); - if (ret) { +static long neuron_copy_from_user(const char *const fname, void * to, const void __user * from, unsigned long n) { + if (copy_from_user(to, from, n)) { pr_err("copy_from_user failed: %s\n", fname); + return -EFAULT; } - return ret; + return 0; +} + +static long neuron_copy_to_user(const char *const fname, void __user * to, const void * from, unsigned long n) { + if (copy_to_user(to, from, n)) { + pr_err("copy_to_user failed: %s\n", fname); + return -EFAULT; + } + return 0; } static int ncdev_ncid_valid(uint32_t nc_id) @@ -134,7 +146,7 @@ static int ncdev_dma_queue_init(struct neuron_device *nd, void *param) ret = neuron_copy_from_user(__func__, &arg, (struct neuron_ioctl_dma_queue_init *)param, sizeof(arg)); if (ret) { - return -EACCES; + return ret; } if (arg.rx_handle) rx_mc = ncdev_mem_handle_to_mem_chunk(nd, arg.rx_handle); @@ -188,7 +200,6 @@ static int ncdev_dma_queue_init_batch(struct neuron_device *nd, void *param) ret = neuron_copy_from_user(__func__, arg, (struct neuron_ioctl_dma_queue_init_batch *)param, sizeof(struct neuron_ioctl_dma_queue_init_batch)); if (ret) { - ret = -EACCES; goto done; } @@ -255,7 +266,7 @@ static int ncdev_dma_copy_descriptors(struct neuron_device *nd, unsigned int cmd } remaining = num_descs * sizeof(union udma_desc); - ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, MAX_DMA_DESC_SIZE, 0, MEM_LOC_HOST, 0, 0, mc->nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &src_mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, MAX_DMA_DESC_SIZE, 0, MEM_LOC_HOST, 0, mc->nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &src_mc); if (ret) { return -ENOMEM; } @@ -424,7 +435,7 @@ static int ncdev_mem_alloc(struct neuron_device *nd, void *param) ret = neuron_copy_from_user(__func__, &mem_alloc_arg, (struct neuron_ioctl_mem_alloc *)param, sizeof(mem_alloc_arg)); if (ret) - return -EACCES; + return ret; mem_alloc_category_t mem_alloc_type; if (mem_alloc_arg.host_memory) { location = MEM_LOC_HOST; @@ -434,8 +445,8 @@ static int ncdev_mem_alloc(struct neuron_device *nd, void *param) location = MEM_LOC_DEVICE; mem_alloc_type = NEURON_MEMALLOC_TYPE_UNKNOWN_DEVICE; } - ret = mc_alloc_align(nd, MC_LIFESPAN_CUR_PROCESS, mem_alloc_arg.size, 0, location, mem_alloc_arg.dram_channel, - mem_alloc_arg.dram_region, mem_alloc_arg.nc_id, mem_alloc_type, &mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_CUR_PROCESS, mem_alloc_arg.size, 0, location, mem_alloc_arg.hbm_index, + mem_alloc_arg.nc_id, mem_alloc_type, &mc); if (ret) return ret; @@ -467,8 +478,7 @@ static int ncdev_mem_alloc_libnrt(struct neuron_device *nd, unsigned int cmd, vo u32 host_memory; u64 size; u64 align; - u32 dram_channel; - u32 dram_region; + u32 hbm_index; u32 nc_id; mem_alloc_category_t mem_type; @@ -484,8 +494,7 @@ static int ncdev_mem_alloc_libnrt(struct neuron_device *nd, unsigned int cmd, vo size = mem_alloc_arg.size; align = mem_alloc_arg.align; host_memory = mem_alloc_arg.host_memory; - dram_channel = mem_alloc_arg.dram_channel; - dram_region = mem_alloc_arg.dram_region; + hbm_index = mem_alloc_arg.hbm_index; nc_id = mem_alloc_arg.nc_id; mem_handle = mem_alloc_arg.mem_handle; if (host_memory) { @@ -503,8 +512,7 @@ static int ncdev_mem_alloc_libnrt(struct neuron_device *nd, unsigned int cmd, vo size = mem_alloc_arg.size; align = mem_alloc_arg.align; host_memory = mem_alloc_arg.host_memory; - dram_channel = mem_alloc_arg.dram_channel; - dram_region = mem_alloc_arg.dram_region; + hbm_index = mem_alloc_arg.hbm_index; nc_id = mem_alloc_arg.nc_id; mem_type = mem_alloc_arg.mem_type; mem_handle = mem_alloc_arg.mem_handle; @@ -518,8 +526,7 @@ static int ncdev_mem_alloc_libnrt(struct neuron_device *nd, unsigned int cmd, vo size = mem_alloc_arg.size; align = mem_alloc_arg.align; host_memory = mem_alloc_arg.host_memory; - dram_channel = mem_alloc_arg.dram_channel; - dram_region = mem_alloc_arg.dram_region; + hbm_index = mem_alloc_arg.hbm_index; nc_id = mem_alloc_arg.nc_id; mem_type = mem_alloc_arg.mem_type; mem_handle = mem_alloc_arg.mem_handle; @@ -534,8 +541,7 @@ static int ncdev_mem_alloc_libnrt(struct neuron_device *nd, unsigned int cmd, vo size = mem_alloc_arg.size; align = mem_alloc_arg.align; host_memory = mem_alloc_arg.host_memory; - dram_channel = mem_alloc_arg.dram_channel; - dram_region = mem_alloc_arg.dram_region; + hbm_index = mem_alloc_arg.hbm_index; nc_id = mem_alloc_arg.nc_id; mem_type = mem_alloc_arg.mem_type; mem_handle = mem_alloc_arg.mem_handle; @@ -544,7 +550,7 @@ static int ncdev_mem_alloc_libnrt(struct neuron_device *nd, unsigned int cmd, vo location = MEM_LOC_HOST; else location = MEM_LOC_DEVICE; - ret = mc_alloc_align(nd, MC_LIFESPAN_CUR_PROCESS, size, align, location, dram_channel, dram_region, nc_id, mem_type, &mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_CUR_PROCESS, size, align, location, hbm_index, nc_id, mem_type, &mc); if (ret) return ret; @@ -578,7 +584,7 @@ static int ncdev_mem_alloc_libnrt(struct neuron_device *nd, unsigned int cmd, vo location = MEM_LOC_HOST; else location = MEM_LOC_DEVICE; - ret = mc_alloc_align(nd, MC_LIFESPAN_CUR_PROCESS, size, align, location, dram_channel, dram_region, nc_id, mem_type, &mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_CUR_PROCESS, size, align, location, hbm_index, nc_id, mem_type, &mc); if (ret) return ret; @@ -723,17 +729,58 @@ static int ncdev_get_dmabuf_fd(void *param) { int ret; struct neuron_ioctl_dmabuf_fd arg; - int dmabuf_fd; + int dmabuf_fd; + u64 offset; ret = neuron_copy_from_user(__func__, &arg, param, sizeof(arg)); if (ret) return ret; - ret = ndmabuf_get_fd(arg.va, arg.size, &dmabuf_fd); + if (!arg.fd) + return -EINVAL; + + if (arg.va & (PAGE_SIZE - 1)) + return -EINVAL; + + ret = ndmabuf_get_fd(arg.va, arg.size, &dmabuf_fd, &offset); if (ret) return ret; - return copy_to_user(arg.fd, &dmabuf_fd, sizeof(dmabuf_fd)); + if (copy_to_user(arg.fd, &dmabuf_fd, sizeof(dmabuf_fd))) + return -EFAULT; + + return 0; +} + +static int ncdev_get_dmabuf_fd_v2(void *param) +{ + int ret; + struct neuron_ioctl_dmabuf_fd_v2 arg; + int dmabuf_fd; + u64 offset; + + ret = neuron_copy_from_user(__func__, &arg, param, sizeof(arg)); + if (ret) + return ret; + + if (!arg.fd || !arg.offset) + return -EINVAL; + + ret = ndmabuf_get_fd(arg.va, arg.size, &dmabuf_fd, &offset); + if (ret) + return ret; + + if (copy_to_user(arg.fd, &dmabuf_fd, sizeof(dmabuf_fd))) + goto err_close_fd; + + if (copy_to_user(arg.offset, &offset, sizeof(offset))) + goto err_close_fd; + + return 0; + +err_close_fd: + close_fd(dmabuf_fd); + return -EFAULT; } static int ncdev_mem_free(struct neuron_device *nd, void *param) @@ -748,7 +795,7 @@ static int ncdev_mem_free(struct neuron_device *nd, void *param) return ret; mc = ncdev_mem_handle_to_mem_chunk(nd, mem_free_arg.mem_handle); if (!mc) - return -EINVAL; + return -ENOENT; trace_ioctl_mem_alloc(nd, mc); mc_free(&mc); return 0; @@ -790,13 +837,13 @@ static int ncdev_memset(struct neuron_device *nd, unsigned int cmd, void *param) mc = ncdev_mem_handle_to_mem_chunk(nd, mem_handle); if (!mc) { - return -EINVAL; + return -ENOENT; } // check access is within the range. if (!mc_access_is_within_bounds(mc, offset, size)) { pr_err("offset+size is too large for mem handle\n"); - return -EINVAL; + return -E2BIG; } ret = ndma_memset(nd, mc, offset, value, size); @@ -847,17 +894,17 @@ static int ncdev_mem_copy(struct neuron_device *nd, unsigned int cmd, void *para src_mc = ncdev_mem_handle_to_mem_chunk(nd, src_mem_handle); dst_mc = ncdev_mem_handle_to_mem_chunk(nd, dst_mem_handle); if (!src_mc || !dst_mc) - return -EINVAL; + return -ENOENT; // check access is within the range. if (!mc_access_is_within_bounds(src_mc, src_offset, size)) { pr_err("src offset+size is too large for mem handle\n"); - return -EINVAL; + return -E2BIG; } // check access is within the range. if (!mc_access_is_within_bounds(dst_mc, dst_offset, size)) { pr_err("dst offset+size is too large for mem handle\n"); - return -EINVAL; + return -E2BIG; } ret = ndma_memcpy_mc(nd, src_mc, dst_mc, src_offset, dst_offset, size); if (ret) { @@ -922,17 +969,17 @@ static int ncdev_mem_copy_async(struct neuron_device *nd, unsigned int cmd, void src_mc = ncdev_mem_handle_to_mem_chunk(nd, src_mem_handle); dst_mc = ncdev_mem_handle_to_mem_chunk(nd, dst_mem_handle); if (!src_mc || !dst_mc) - return -EINVAL; + return -ENOENT; // check access is within the range. if (!mc_access_is_within_bounds(src_mc, src_offset, size)) { pr_err("src offset+size is too large for mem handle\n"); - return -EINVAL; + return -E2BIG; } // check access is within the range. if (!mc_access_is_within_bounds(dst_mc, dst_offset, size)) { pr_err("dst offset+size is too large for mem handle\n"); - return -EINVAL; + return -E2BIG; } ret = ndma_memcpy_mc_async(nd, src_mc, dst_mc, src_offset, dst_offset, size, host_prefetch_addr, pwait_handle, wait_handle); @@ -942,10 +989,12 @@ static int ncdev_mem_copy_async(struct neuron_device *nd, unsigned int cmd, void } // return the new wait handle - ret = copy_to_user((void *)param, &arg, arg_size); + ret = neuron_copy_to_user(__func__, (void *)param, &arg, arg_size); + if (ret) + return ret; trace_ioctl_mem_copy(nd, src_mc, dst_mc); - return ret; + return 0; } static int ncdev_mem_copy_async_wait(struct neuron_device *nd, void *param) @@ -963,7 +1012,7 @@ static int ncdev_mem_copy_async_wait(struct neuron_device *nd, void *param) dst_mc = ncdev_mem_handle_to_mem_chunk(nd, arg.dst_mem_handle); if (!src_mc || !dst_mc) { pr_err("dma memcpy wait failed. invalid mem chunk handle\n"); - return -EINVAL; + return -ENOENT; } if ((arg.pwait_handle < NEURON_DMA_H2T_CTX_HANDLE_ASYNC1) || (arg.pwait_handle > NEURON_DMA_H2T_CTX_HANDLE_ASYNC2)) { @@ -1007,7 +1056,7 @@ static int ncdev_program_engine(struct neuron_device *nd, void *param) if (ncdev_verify_mem_region(arg.dst + arg.offset, arg.size)) return -ENOMEM; - ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, arg.size, 0, MEM_LOC_HOST, 0, 0, 0, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &src_mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, arg.size, 0, MEM_LOC_HOST, 0, 0, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &src_mc); if (ret) { ret = -ENOMEM; return ret; @@ -1068,7 +1117,7 @@ static int ncdev_program_engine_nc(struct neuron_device *nd, unsigned int cmd, v if (ncdev_verify_mem_region(dst + offset, size)) return -ENOMEM; - ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, size, 0, MEM_LOC_HOST, 0, 0, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &src_mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, size, 0, MEM_LOC_HOST, 0, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &src_mc); if (ret) { pr_err("engine programming dma mc_alloc_align failed. nc_id: %d addr: %llu size: %llu err: %d\n", nc_id, dst + offset, size, ret); ret = -ENOMEM; @@ -1128,11 +1177,11 @@ static int ncdev_mem_buf_copy(struct neuron_device *nd, unsigned int cmd, void * mc = ncdev_mem_handle_to_mem_chunk(nd, mem_handle); if (!mc) - return -EINVAL; + return -ENOENT; // check access is within the range. if (!mc_access_is_within_bounds(mc, offset, size)) { pr_err("offset+size is too large for mem handle\n"); - return -EINVAL; + return -E2BIG; } if (copy_to_mem_handle) @@ -1144,7 +1193,7 @@ static int ncdev_mem_buf_copy(struct neuron_device *nd, unsigned int cmd, void * if (copy_to_mem_handle) { ret = neuron_copy_from_user(__func__, mc->va + offset, buffer, size); } else { - ret = copy_to_user(buffer, mc->va + offset, size); + ret = neuron_copy_to_user(__func__, buffer, mc->va + offset, size); } return ret; } else { @@ -1153,7 +1202,7 @@ static int ncdev_mem_buf_copy(struct neuron_device *nd, unsigned int cmd, void * u32 copy_offset = 0; u32 remaining = size; u32 copy_size = 0; - ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, MAX_DMA_DESC_SIZE, 0, MEM_LOC_HOST, 0, 0, + ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, MAX_DMA_DESC_SIZE, 0, MEM_LOC_HOST, 0, mc->nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &src_mc); if (ret) { ret = -ENOMEM; @@ -1177,7 +1226,7 @@ static int ncdev_mem_buf_copy(struct neuron_device *nd, unsigned int cmd, void * if (ret) { break; } - ret = copy_to_user(buffer + copy_offset, src_mc->va, copy_size); + ret = neuron_copy_to_user(__func__, buffer + copy_offset, src_mc->va, copy_size); if (ret) { break; } @@ -1197,6 +1246,8 @@ static int ncdev_mem_buf_copy(struct neuron_device *nd, unsigned int cmd, void * #define BAR4_WR_THRESHOLD_MAX (PAGE_SIZE*2) static int ncdev_mem_buf_zerocopy64(struct neuron_device *nd, unsigned int cmd, void *param) { + u64 sequence_num; + void *context; void *buffer; struct mem_chunk *mc; u64 mem_handle; @@ -1209,13 +1260,14 @@ static int ncdev_mem_buf_zerocopy64(struct neuron_device *nd, unsigned int cmd, struct neuron_ioctl_mem_buf_copy64zc arg; bool use_bar4_wr; - // TODO remove at some point + // Preserve ABI safety: reject legacy/new struct size mismatches explicitly + // before copying userspace memory into the ioctl. if (_IOC_SIZE(cmd) != sizeof(arg)) { - pr_err_once("error experimental zerocopy API is now obsolete. Please upgrade to latest driver"); - return -EINVAL; + pr_err_once("The zerocopy API version is obsolete. Please upgrade to the latest driver"); + return -EINVAL; } - ret = neuron_copy_from_user(__func__, &arg, (struct neuron_ioctl_mem_buf_copy64 *)param, sizeof(arg)); + ret = neuron_copy_from_user(__func__, &arg, (struct neuron_ioctl_mem_buf_copy64zc *)param, sizeof(arg)); if (ret) return ret; mem_handle = arg.mem_handle; @@ -1224,6 +1276,8 @@ static int ncdev_mem_buf_zerocopy64(struct neuron_device *nd, unsigned int cmd, offset = arg.offset; size = arg.size; h2t_qid = arg.h2t_qid; + sequence_num = arg.sequence_num; + context = arg.context; mc = ncdev_mem_handle_to_mem_chunk(nd, mem_handle); if (!mc) @@ -1286,7 +1340,9 @@ static int ncdev_mem_buf_zerocopy64(struct neuron_device *nd, unsigned int cmd, op.buffer = buffer; op.size = size; - ret = ndma_zerocopy_submit(nd, nc_id, &op, 1, dev_base, qid, copy_to_mem_handle ? true : false, 0); + ret = ndma_zerocopy_submit(nd, nc_id, &op, 1, dev_base, qid, + copy_to_mem_handle ? true : false, + sequence_num, context); } return ret; @@ -1466,7 +1522,7 @@ static int ncdev_mem_buf_zerocopy64_batch(struct neuron_device *nd, void *param) } // use the zero-copy batch function for ops within a single batch - ret = ndma_zerocopy_submit(nd, nc_id, batch->ops_ptr, batch->num_ops, dev_base, qid, arg.is_copy_to_device, arg.sequence_num); + ret = ndma_zerocopy_submit(nd, nc_id, batch->ops_ptr, batch->num_ops, dev_base, qid, arg.is_copy_to_device, arg.sequence_num, NULL); if (ret) { pr_err("batch zero-copy DMA failed on batch %d on nd%02d: %d\n", i, nd->device_index, ret); goto cleanup; @@ -1970,6 +2026,8 @@ static long ncdev_driver_info(unsigned int cmd, void *param) NEURON_DRIVER_FEATURE_MEM_ALLOC64 | NEURON_DRIVER_FEATURE_CONTIGUOUS_SCRATCHPAD | NEURON_DRIVER_FEATURE_ZEROCOPY | NEURON_DRIVER_FEATURE_PINNED_HOST_MEM | NEURON_DRIVER_FEATURE_ALLOC_WITH_PA; + if (ndma_zerocopy_supported()) + driver_info.feature_flags1 |= NEURON_DRIVER_FEATURE_ASYNC_IO; return copy_to_user(param, &driver_info, sizeof(driver_info)); } @@ -2059,7 +2117,7 @@ static long ncdev_nc_nq_init_deprecated(struct neuron_device *nd, void *param) if (ret) return ret; - ret = nnq_init(nd, arg.nc_id, arg.engine_index, arg.nq_type, arg.size, true, 0, 0, + ret = nnq_init(nd, arg.nc_id, arg.engine_index, arg.nq_type, arg.size, true, 0, false, &mc, &arg.mmap_offset); if (ret) return ret; @@ -2079,11 +2137,11 @@ static long ncdev_nc_nq_init_libnrt(struct neuron_device *nd, void *param) if (arg.nq_dev_type == NQ_DEVICE_TYPE_NEURON_CORE) { ret = nnq_init(nd, arg.nq_dev_id, arg.engine_index, arg.nq_type, arg.size, - arg.on_host_memory, arg.dram_channel, arg.dram_region, + arg.on_host_memory, arg.hbm_index, false, &mc, &arg.mmap_offset); } else if (arg.nq_dev_type == NQ_DEVICE_TYPE_TOPSP) { ret = ndhal->ndhal_topsp.ts_nq_init(nd, arg.nq_dev_id, arg.engine_index, arg.nq_type, arg.size, - arg.on_host_memory, arg.dram_channel, arg.dram_region, + arg.on_host_memory, arg.hbm_index, false, &mc, &arg.mmap_offset); } else { return -ENOSYS; @@ -2110,11 +2168,11 @@ static long ncdev_nc_nq_init_with_realloc_libnrt(struct neuron_device *nd, void if (arg.nq_dev_type == NQ_DEVICE_TYPE_NEURON_CORE) { ret = nnq_init(nd, arg.nq_dev_id, arg.engine_index, arg.nq_type, arg.size, - arg.on_host_memory, arg.dram_channel, arg.dram_region, + arg.on_host_memory, arg.hbm_index, arg.force_alloc_mem, &mc, &arg.mmap_offset); } else if (arg.nq_dev_type == NQ_DEVICE_TYPE_TOPSP) { ret = ndhal->ndhal_topsp.ts_nq_init(nd, arg.nq_dev_id, arg.engine_index, arg.nq_type, arg.size, - arg.on_host_memory, arg.dram_channel, arg.dram_region, + arg.on_host_memory, arg.hbm_index, arg.force_alloc_mem, &mc, &arg.mmap_offset); } else { return -ENOSYS; @@ -2227,11 +2285,11 @@ static long ncdev_crwl_nc_range_mark(struct file *filep, unsigned int cmd, void ncd = filep->private_data; if (ncd == NULL) { - return -EINVAL; + return -ENXIO; } nd = ncd->ndev; if (nd == NULL) { - return -EINVAL; + return -ENXIO; } // verify the bitmap is large enough to hold all cores (at compile time) @@ -2242,7 +2300,7 @@ static long ncdev_crwl_nc_range_mark(struct file *filep, unsigned int cmd, void if (size == sizeof(struct neuron_ioctl_crwl_nc_map *)) { size = sizeof(struct neuron_ioctl_crwl_nc_map); } else if (size != sizeof(struct neuron_ioctl_crwl_nc_map_ext)) { - return -EINVAL; + return -ENXIO; } ret = neuron_copy_from_user(__func__, &arg, param, size); @@ -2558,7 +2616,7 @@ static long ncdev_hbm_scrub_start(struct neuron_device *nd, void *param) { if (ret) { return ret; } - if (arg.hbm_index >= ndhal->ndhal_address_map.dram_channels) { + if (arg.hbm_index >= ndhal->ndhal_address_map.num_hbms) { pr_err("HBM scrub: invalid HBM index %d\n", arg.hbm_index); return -1; } @@ -2591,7 +2649,7 @@ static long ncdev_hbm_scrub_start(struct neuron_device *nd, void *param) { completion_bufs[i] = NULL; } int completion_buf_mc_size = DMA_COMPLETION_MARKER_SIZE * 2 * num_dma_engines; - ret = mc_alloc_align(nd, MC_LIFESPAN_CUR_PROCESS, completion_buf_mc_size, 0, MEM_LOC_HOST, 0, 0, arg.nc_id, NEURON_MEMALLOC_TYPE_CODE_HOST, completion_mc_ptr); + ret = mc_alloc_align(nd, MC_LIFESPAN_CUR_PROCESS, completion_buf_mc_size, 0, MEM_LOC_HOST, 0, arg.nc_id, NEURON_MEMALLOC_TYPE_CODE_HOST, completion_mc_ptr); if (ret) { goto scrub_init_fail; } @@ -2617,15 +2675,15 @@ static long ncdev_hbm_scrub_start(struct neuron_device *nd, void *param) { for (i = 0; i= ndhal->ndhal_address_map.dram_channels) { + if (arg.hbm_index >= ndhal->ndhal_address_map.num_hbms) { pr_err("HBM scrub: invalid HBM index %d\n", arg.hbm_index); return -1; } @@ -3015,7 +3073,9 @@ static int ncdev_h2t_dma_alloc_queues(struct neuron_device *nd, unsigned int cmd arg.copy_default_queue = ndhal->ndhal_ndmar.ndmar_get_h2t_def_qid(arg.nc_id); - ret = copy_to_user(param, &arg, sizeof(arg)); + ret = neuron_copy_to_user(__func__, param, &arg, sizeof(arg)); + if (ret) + goto done; done: if (ret) { @@ -3251,6 +3311,8 @@ inline static long ncdev_misc_ioctl(struct file *filep, unsigned int cmd, unsign * over all devices in the user space */ return ncdev_get_dmabuf_fd((void *)param); + } else if (cmd == NEURON_IOCTL_DMABUF_FD_V2) { + return ncdev_get_dmabuf_fd_v2((void *)param); } else if (_IOC_NR(cmd) == _IOC_NR(NEURON_IOCTL_DRIVER_INFO_GET)) { return ncdev_driver_info(cmd, (void*)param); } else if (cmd == NEURON_IOCTL_PRINTK) { @@ -4098,6 +4160,15 @@ static const struct ncdev_class_attr ncdev_class_attrs_pds[] = { NCDEV_CLASS_ATTR(current_perf_profile, ncdev_class_cur_perf_profile_show, NEURON_PLATFORM_TYPE_STD, 0), }; +static const struct ncdev_class_attr ncdev_class_attrs_max[] = { + NCDEV_CLASS_ATTR(node_id, ncdev_class_node_id_show, NEURON_PLATFORM_TYPE_MAX, 0), + NCDEV_CLASS_ATTR(node_cnt, ncdev_class_node_cnt_show, NEURON_PLATFORM_TYPE_MAX, 0), + NCDEV_CLASS_ATTR(reservation_id, ncdev_class_reservation_id_show, NEURON_PLATFORM_TYPE_MAX, 0), + NCDEV_CLASS_ATTR(ultraserver_mode, ncdev_class_ultraserver_mode_show, NEURON_PLATFORM_TYPE_MAX, 0), + NCDEV_CLASS_ATTR(hbm_7200_capable, ncdev_class_hbm_7200_show, NEURON_PLATFORM_TYPE_STD, 0), + NCDEV_CLASS_ATTR(current_perf_profile, ncdev_class_cur_perf_profile_show, NEURON_PLATFORM_TYPE_STD, 0), +}; + static const struct class_attribute class_attr_node_id = __ATTR(node_id, S_IRUGO, ncdev_class_node_id_show, NULL); @@ -4117,6 +4188,7 @@ static const struct { {ncdev_class_attrs, sizeof(ncdev_class_attrs) / sizeof(*ncdev_class_attrs), NEURON_PLATFORM_TYPE_STD}, {ncdev_class_attrs_us, sizeof(ncdev_class_attrs_us) / sizeof(*ncdev_class_attrs_us), NEURON_PLATFORM_TYPE_ULTRASERVER}, {ncdev_class_attrs_pds, sizeof(ncdev_class_attrs_pds) / sizeof(*ncdev_class_attrs_pds), NEURON_PLATFORM_TYPE_PDS}, + {ncdev_class_attrs_max, sizeof(ncdev_class_attrs_max) / sizeof(*ncdev_class_attrs_max), NEURON_PLATFORM_TYPE_MAX}, {NULL, 0, NEURON_PLATFORM_TYPE_INVALID}}; int ncdev_class_attr_init(void) diff --git a/neuron_core.h b/neuron_core.h index dd62669..a2909c3 100644 --- a/neuron_core.h +++ b/neuron_core.h @@ -86,8 +86,6 @@ int nc_event_set(struct neuron_device *nd, u8 nc_id, u16 event_index, u32 value) #define MAX_NQ_TYPE 6 //for v1 4 and v2 6 #define MAX_NQ_ENGINE 16 // for v1 4 engines for v2 16 queues -#define MAX_NQ_SUPPORTED (MAX_NQ_TYPE * MAX_NQ_ENGINE) - /** * nc_get_nq_mem_handle() - Get notification queue's mem handle for given neuron core. * diff --git a/neuron_crwl.c b/neuron_crwl.c index f528f1c..b89f302 100644 --- a/neuron_crwl.c +++ b/neuron_crwl.c @@ -193,7 +193,7 @@ int ncrwl_nc_range_mark(u32 nc_count, u32 start_nc_index, u32 end_nc_index, if (start_nc_index > end_nc_index || start_nc_index >= MAX_NEURON_DEVICE_COUNT * MAX_NC_PER_DEVICE || end_nc_index >= MAX_NEURON_DEVICE_COUNT * MAX_NC_PER_DEVICE) - return -EINVAL; + return -ENXIO; mutex_lock(&ncrwl_range_lock); for (i = start_nc_index; i <= end_nc_index; i++) { int range_len = 1; diff --git a/neuron_device.h b/neuron_device.h index fa67c5b..b6a9190 100644 --- a/neuron_device.h +++ b/neuron_device.h @@ -63,7 +63,7 @@ enum neuron_device_state { struct neuron_hbm_scrub_ctx { struct mem_chunk *rx_mc[NUM_DMA_ENG_PER_DEVICE]; struct mem_chunk *tx_mc[NUM_DMA_ENG_PER_DEVICE]; - struct mem_chunk *completion_marker_buf[MAX_DRAM_CHANNELS]; // one memchunk shared by all DMA engines for an HBM to reduce internal fragmentation + struct mem_chunk *completion_marker_buf[MAX_NUM_HBMS]; // one memchunk shared by all DMA engines for an HBM to reduce internal fragmentation struct mem_chunk *hostbuf_mc[NUM_DMA_ENG_PER_DEVICE]; }; @@ -89,9 +89,9 @@ struct neuron_device { struct neuron_mempool_set mpset; // memory chunk allocated for notification queue in each neuron core. - struct mem_chunk *nq_mc[MAX_NC_PER_DEVICE][MAX_NQ_SUPPORTED]; + struct mem_chunk **nq_mc[MAX_NC_PER_DEVICE]; // memory chunk allocated for notification queue in each TOP_SP. - struct mem_chunk *ts_nq_mc[MAX_TS_PER_DEVICE][MAX_NQ_SUPPORTED]; + struct mem_chunk **ts_nq_mc[MAX_TS_PER_DEVICE]; // memory chunk for setting device mem struct mem_chunk *memset_mc; diff --git a/neuron_dhal.h b/neuron_dhal.h index be38ead..71548b5 100644 --- a/neuron_dhal.h +++ b/neuron_dhal.h @@ -48,7 +48,8 @@ struct ndhal_address_map { int dma_eng_per_nc; int seng_dma_eng_per_nd; int h2d_dma_eng_per_nd; - int dram_channels; + int num_hbms; + u16 nq_per_nc; }; struct ndhal_reset { @@ -62,7 +63,7 @@ struct ndhal_reset { struct ndhal_topsp { int (*ts_nq_init) (struct neuron_device *nd, u8 ts_id, u8 eng_index, u32 nq_type, u32 size, - u32 on_host_memory, u32 dram_channel, u32 dram_region, + u32 on_host_memory, u32 hbm_index, bool force_alloc_mem, struct mem_chunk **nq_mc, u64 *mmap_offset); void (*ts_nq_destroy_one) (struct neuron_device *nd, u8 ts_id); u8 (*ts_nq_get_nqid)(struct neuron_device *nd, u8 index, u32 nq_type); @@ -81,8 +82,8 @@ struct ndhal_nq { struct ndhal_mpset { int mp_min_alloc_size; - u64 device_dram_effective_base_addr[MAX_DRAM_CHANNELS]; - u64 device_dram_end_addr[MAX_DRAM_CHANNELS]; + u64 device_dram_effective_base_addr[MAX_NUM_HBMS]; + u64 device_dram_end_addr[MAX_NUM_HBMS]; bool small_pool_supported; void (*mpset_set_dram_and_mpset_info) (struct neuron_mempool_set *mpset, u64 *device_dram_addr, u64 *device_dram_size); }; diff --git a/neuron_dma.c b/neuron_dma.c index 901048b..fc768ac 100644 --- a/neuron_dma.c +++ b/neuron_dma.c @@ -979,6 +979,7 @@ struct ndma_h2t_zcdma_context { // host memory buffer which driver polls on for completions; // managed by completion_pool in ctx queue u64 sequence_num; // async sequence number; 0 for sync transfers + void *context; // async completion context // Async-only struct mm_struct *mm; // mm that owns the user buffers @@ -1015,6 +1016,7 @@ static void ndma_zc_release_ctx(struct ndma_h2t_zcdma_context *ctx, u64 *nr_pinn ctx->state = NDMA_INVALID; ctx->sequence_num = 0; + ctx->context = NULL; } /* H2D DMA Completion Queue (CQ) */ @@ -1028,7 +1030,7 @@ int ndma_h2d_compl_queue_init(struct neuron_device *nd, struct ndma_h2d_compl_qu queue_size = sizeof(neuron_h2d_dma_compl_queue_t) + (NDMA_H2D_COMPL_QUEUE_CAPACITY * sizeof(neuron_h2d_dma_compl_queue_entry_t)); ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, queue_size, 0, - MEM_LOC_HOST, 0, 0, 0, + MEM_LOC_HOST, 0, 0, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &mc); if (ret) { pr_err("failed to allocate h2d dma completion queue mc: %d\n", ret); @@ -1263,19 +1265,21 @@ static struct ndma_h2t_zcdma_context *ndma_ctx_queue_pop_submitted(struct ndma_c /* Failure-path helper. * Given a sequence number of a async request, wait for any matching submitted ctxs, then reset all matching ctxs. * This prevents further remote pinning and submitting on a failed async request. - * Mostly used in failure and cleanup paths, so don't stop on failed DMAs. + * Once a wait times out, skip waiting on subsequent contexts (engine is stuck). */ static void ndma_ctx_queue_drain_sequence(struct ndma_ctx_queue *queue, u64 sequence_num) { u32 idx; + bool timed_out = false; for (idx = queue->head; idx != queue->tail; idx = (idx + 1) & queue->capacity_mask) { struct ndma_h2t_zcdma_context *ctx = &queue->entries[idx]; if (ctx->sequence_num == sequence_num) { - // wait for already submitted DMAs to complete. - if (ctx->state == NDMA_SUBMITTED) { - ndma_memcpy_wait_for_completion(ctx->eng, ctx->ring, ctx->nr_desc + 1, ctx->completion_ptr, false, false); + // wait for already submitted DMAs to complete, unless engine already timed out. + if (ctx->state == NDMA_SUBMITTED && !timed_out) { + if (ndma_memcpy_wait_for_completion(ctx->eng, ctx->ring, ctx->nr_desc + 1, ctx->completion_ptr, false, false)) + timed_out = true; } // release pinned pages and mm, and set state to invalid (tombstone). @@ -1292,17 +1296,20 @@ static void ndma_ctx_queue_drain_sequence(struct ndma_ctx_queue *queue, u64 sequ /* Failure-path helper. * Wait for submitted contexts from head up to (but not including) first_pinned_unsubmitted. * Unpin from head up to (but not including) first_unpinned. - * Mostly used in failure and cleanup paths, so don't stop on failed DMAs. + * Once a wait times out, skip waiting on subsequent contexts (engine is stuck). */ static void ndma_ctx_queue_drain(struct ndma_eng *eng, struct ndma_ring *ring, struct ndma_ctx_queue *queue) { + bool timed_out = false; + while (!ndma_ctx_queue_is_empty(queue)) { struct ndma_h2t_zcdma_context *ctx = ndma_ctx_queue_pop_head(queue); - if (ctx->state == NDMA_SUBMITTED) { - ndma_memcpy_wait_for_completion(eng, ring, ctx->nr_desc + 1, ctx->completion_ptr, false, false); + if (ctx->state == NDMA_SUBMITTED && !timed_out) { + if (ndma_memcpy_wait_for_completion(eng, ring, ctx->nr_desc + 1, ctx->completion_ptr, false, false)) + timed_out = true; } ndma_zc_release_ctx(ctx, &queue->nr_pinned_pages); @@ -1670,7 +1677,8 @@ int ndma_zerocopy_submit(struct neuron_device *nd, dma_addr_t dev_base, int qid, bool direction, - u64 sequence_num) + u64 sequence_num, + void *context) { int ret = 0; int i = 0; @@ -1757,6 +1765,7 @@ int ndma_zerocopy_submit(struct neuron_device *nd, cur_ctx->mm = NULL; cur_ctx->pid = task_tgid_nr(current); cur_ctx->sequence_num = sequence_num; + cur_ctx->context = context; /* Pin now if possible; otherwise capture mm for remote pinning (async only). */ if (can_pin) { @@ -1892,10 +1901,10 @@ static int ndma_zerocopy_complete(struct neuron_device *nd, if (ret) { err = ret; pr_err("async h2d dma completion failed for seq num %llu: %d\n", submitted_ctx->sequence_num, ret); - ndma_h2d_compl_queue_put(&ring->dma_compl_queue, submitted_ctx->sequence_num, ret, NULL); + ndma_h2d_compl_queue_put(&ring->dma_compl_queue, submitted_ctx->sequence_num, ret, submitted_ctx->context); ndma_ctx_queue_drain_sequence(ctx_queue, submitted_ctx->sequence_num); } else if (submitted_ctx->last) { - ndma_h2d_compl_queue_put(&ring->dma_compl_queue, submitted_ctx->sequence_num, 0, NULL); + ndma_h2d_compl_queue_put(&ring->dma_compl_queue, submitted_ctx->sequence_num, 0, submitted_ctx->context); } ndma_zc_release_ctx(submitted_ctx, &ctx_queue->nr_pinned_pages); @@ -1914,7 +1923,7 @@ static int ndma_zerocopy_complete(struct neuron_device *nd, if (ret) { err = ret; pr_err("async h2d dma submission failed for seq num %llu: %d\n", pinned_unsubmitted_ctx->sequence_num, ret); - ndma_h2d_compl_queue_put(&ring->dma_compl_queue, pinned_unsubmitted_ctx->sequence_num, ret, NULL); + ndma_h2d_compl_queue_put(&ring->dma_compl_queue, pinned_unsubmitted_ctx->sequence_num, ret, pinned_unsubmitted_ctx->context); ndma_ctx_queue_drain_sequence(ctx_queue, pinned_unsubmitted_ctx->sequence_num); } else { ndma_ctx_queue_inc_first_pinned_unsubmitted(ctx_queue); @@ -1934,7 +1943,7 @@ static int ndma_zerocopy_complete(struct neuron_device *nd, if (ret) { err = ret; pr_err("async h2d dma remote pinning failed for seq num %llu: %d\n", unpinned_ctx->sequence_num, ret); - ndma_h2d_compl_queue_put(&ring->dma_compl_queue, unpinned_ctx->sequence_num, ret, NULL); + ndma_h2d_compl_queue_put(&ring->dma_compl_queue, unpinned_ctx->sequence_num, ret, unpinned_ctx->context); ndma_ctx_queue_drain_sequence(ctx_queue, unpinned_ctx->sequence_num); } else { ndma_ctx_queue_inc_first_unpinned(ctx_queue); diff --git a/neuron_dma.h b/neuron_dma.h index 0661a49..787e8f8 100644 --- a/neuron_dma.h +++ b/neuron_dma.h @@ -211,6 +211,7 @@ bool ndma_zerocopy_supported(void); * @qid: Queue identifier to submit descriptors on. * @direction: true for host-to-device, false for device-to-host. * @sequence_num: sequence number under async submission; 0 for sync. + * @context: opaque context pointer returned in the completion queue for async submissions. * * DMA data between a user space virtual address range and a contiguous location in device memory. * In order to do this, we need to know the physical pages are associated with @@ -260,7 +261,8 @@ int ndma_zerocopy_submit(struct neuron_device *nd, dma_addr_t dev_base, int qid, bool direction, - u64 sequence_num); + u64 sequence_num, + void *context); /** * Pre-pinned host memory support diff --git a/neuron_dmabuf.c b/neuron_dmabuf.c index 2f53a07..57dc738 100644 --- a/neuron_dmabuf.c +++ b/neuron_dmabuf.c @@ -8,6 +8,8 @@ #include "neuron_mmap.h" #include "neuron_pci.h" +#include + /* Below is a short description of how dma-buf works (with libfabric and EFA driver) * - Memory Registration * - Libfabric (user-space) requests a dmabuf file-descriptor (FD) @@ -19,7 +21,6 @@ * - Memory Deregistration * - These APIs get invoked in Neuron driver: * - ndmabuf_unmap - * - ndmabuf_detach * - ndmabuf_release * * See internal documentation: Moving-to-dma-buf-for-EFA-Memory-Registration for more information. @@ -27,6 +28,8 @@ #if LINUX_VERSION_CODE >= KERNEL_VERSION(5, 15, 0) +#define SG_MAX_LEN ALIGN_DOWN((u64)U32_MAX, PAGE_SIZE) + #if LINUX_VERSION_CODE >= KERNEL_VERSION(5, 16, 0) #if (LINUX_VERSION_CODE >= KERNEL_VERSION(6, 13, 0)) || \ (defined(RHEL_RELEASE_CODE) && (RHEL_RELEASE_CODE >= RHEL_RELEASE_VERSION(10, 0))) @@ -41,98 +44,33 @@ struct ndmabuf_private_data { /* Virtual address in userspace */ void *va; - /* size of the buffer */ - u64 size; - - /* Set to 1 when an external driver is attached to dmabuf object */ - bool is_attached; - - /* File-descriptor associated with the dmabuf object */ - int fd; + /* Page-aligned virtual address */ + void *aligned_va; /* Neuron device index associated with the dmabuf object */ int device_idx; }; /* Invoked when an external driver is being attached to a dmabuf object */ -static int ndmabuf_attach( - struct dma_buf * dmabuf, - struct dma_buf_attachment *attachment) { - struct ndmabuf_private_data *private_data; - - private_data = dmabuf->priv; - if (private_data == NULL) { - pr_err("ndmabuf_attach: Neuron context (private data) in dmabuf was freed prematurely!"); - return -EINVAL; - } - - if (!__sync_bool_compare_and_swap(&private_data->is_attached, 0, 1)) { - /* An external driver has to detach before calling attach again */ - pr_err("ndmabuf_attach: Only one device is allowed to be attached to a Neuron dmabuf object\n"); - return -EPERM; - } - +static int ndmabuf_attach(struct dma_buf * dmabuf, struct dma_buf_attachment *attachment) +{ return 0; } -/* Invoked when an external driver is being detached from a dmabuf object */ -static void ndmabuf_detach( - struct dma_buf * dmabuf, - struct dma_buf_attachment *attachment) { - struct ndmabuf_private_data *private_data; - struct neuron_device *nd; - - private_data = dmabuf->priv; - if (private_data == NULL) { - pr_err("ndmabuf_detach: Neuron context (private data) in dmabuf was freed prematurely!"); - return; - } - - if (!__sync_bool_compare_and_swap(&private_data->is_attached, 1, 0)) { - pr_err("ndmabuf_detach: dmabuf object is already detached, " - "multiple detach calls are not allowed for the same dmabuf object\n"); - return; - } - - nd = neuron_pci_get_device(private_data->device_idx); - if (nd == NULL) { - pr_err("ndmabuf_detach: Failed to retrieve nd%d, is the device closed?\n", - private_data->device_idx); - BUG_ON(true); /* Very bad news - no point in moving further */ - } - - /* By design, dma-buf doesn't release its FD until the - * process exits. That might create a problem (run out of FDs) - * for Neuron applications that load/unload multiple NEFFs - * within a single process. So, forcefuly mark the FD as - * unused to recycle it. Also, put back the associated file. - */ - /* Special : Make sure the process is still attached to the device. - * This is to cover the case where a process dies and cleans up - * FD before detach is called in the dma-buf flow. - * For example, when an application dies between memory registration and - * de-registration, FD is already uninstalled by the process before - * de-registration is called in the kernel */ - if (npid_is_attached(nd)) { - put_unused_fd(private_data->fd); - fput(dmabuf->file); - } - - return; -} - /* Invoked when an external driver wants to retrieve pages * (physical addresses) of a Neuron device buffer */ -static struct sg_table *ndmabuf_map( - struct dma_buf_attachment *attachment, - enum dma_data_direction dir) { - struct dma_buf *dmabuf; +static struct sg_table *ndmabuf_map(struct dma_buf_attachment *attachment, enum dma_data_direction dir) +{ struct ndmabuf_private_data *private_data; + struct neuron_device *nd; + struct nmmap_node *mmap; struct scatterlist *sg; + struct dma_buf *dmabuf; struct sg_table *sgt; - int sg_entries, sg_idx, ret; - u64 pa, size_aligned; - u32 page_size; + u64 pa, remaining; + int sg_entries; + int sg_idx; + int ret; dmabuf = attachment->dmabuf; private_data = dmabuf->priv; @@ -141,31 +79,34 @@ static struct sg_table *ndmabuf_map( return ERR_PTR(-EINVAL); } - /* NOTE: Theoritically, __sync builtin is not required here but - * just being paranoid in case the order of dmabuf function calls - * gets messed up outside of Neuron driver. */ - if (!__sync_bool_compare_and_swap(&private_data->is_attached, 1, 1)) { - pr_err("ndmabuf_map: Must attach() before map()\n"); - return ERR_PTR(-EPERM); - } - - /* Single entry in scatterlist */ - page_size = PAGE_SIZE; - /* Find the matching mmap node in the device. * Populate sg_table using the information in mmap. * Also, store some context to the private data inside dmabuf object. */ - - struct neuron_device *nd = neuron_pci_get_device(private_data->device_idx); + nd = neuron_pci_get_device(private_data->device_idx); if (nd == NULL) { - pr_err("ndmabuf_detach: Failed to retrieve nd%d, is the device closed?\n", + pr_err("ndmabuf_map: Failed to retrieve nd%d, is the device closed?\n", private_data->device_idx); return ERR_PTR(-EINVAL); } + /* Device memory is physically contiguous within an mmap region, so use + * max-sized SG entries (capped at U32_MAX-PAGE_SIZE due to the 32-bit + * scatterlist length field). Page size optimization will be handled by + * EFA driver via IB subsystem's ib_umem_find_best_pgsz() */ + sg_entries = DIV_ROUND_UP_ULL(dmabuf->size, SG_MAX_LEN); + + sgt = kzalloc(sizeof(*sgt), GFP_KERNEL); + if (!sgt) + return ERR_PTR(-ENOMEM); + + ret = sg_alloc_table(sgt, sg_entries, GFP_KERNEL); + if (ret) { + goto err_free_sgt; + } + write_lock(&nd->mpset.rbmmaplock); - struct nmmap_node *mmap = nmmap_search_va(nd, private_data->va); + mmap = nmmap_search_va(nd, private_data->va); if (mmap == NULL) { pr_err("ndmabuf_map: mmap node (nd:%d va:0x%llx) was freed prematurely!\n", private_data->device_idx, (u64)private_data->va); @@ -173,37 +114,19 @@ static struct sg_table *ndmabuf_map( goto err_unlock; } - sgt = kzalloc(sizeof(*sgt), GFP_KERNEL); - if (!sgt) { - ret = -ENOMEM; - goto err_unlock; - } - - size_aligned = ALIGN(dmabuf->size, page_size); - sg_entries = size_aligned/page_size; - ret = sg_alloc_table(sgt, sg_entries, GFP_KERNEL | __GFP_ZERO); - if (ret) { - goto err_sgt_free; - } + pa = mmap->pa + (private_data->aligned_va - mmap->va); - pa = mmap->pa + (private_data->va - mmap->va); - - /* Memory allocated & mapped using driver are always PAGE_SIZE aligned. - * Make sure the pa we get is page size aligned */ - if (pa % PAGE_SIZE != 0) { - pr_err("physical address is not %ld aligned for pid:%d", PAGE_SIZE, task_tgid_nr(current)); - ret = -EINVAL; - goto err_sgt_free; - } - - /* Populate the sg_table */ + remaining = dmabuf->size; for_each_sgtable_dma_sg(sgt, sg, sg_idx) { - sg_set_page(sg, NULL, page_size, 0); + u32 len = min_t(u64, remaining, SG_MAX_LEN); sg_dma_address(sg) = pa; - sg_dma_len(sg) = page_size; - pa += page_size; + sg_dma_len(sg) = len; + pa += len; + remaining -= len; } + WARN_ON(remaining != 0); + /* Increment the usage count */ mmap->dmabuf_ref_cnt++; @@ -211,11 +134,13 @@ static struct sg_table *ndmabuf_map( return sgt; -err_sgt_free: - kfree(sgt); - err_unlock: write_unlock(&nd->mpset.rbmmaplock); + +err_free_sgt: + sg_free_table(sgt); + kfree(sgt); + return ERR_PTR(ret); } @@ -232,14 +157,6 @@ static void ndmabuf_unmap( return; } - /* NOTE: Theoritically, __sync builtin is not required here but - * just being paranoid in case the order of dmabuf function calls - * gets messed up outside of Neuron driver. */ - if (!__sync_bool_compare_and_swap(&private_data->is_attached, 1, 1)) { - pr_err("ndmabuf_unmap: Must attach() before unmap()\n"); - BUG_ON(true); /* Very bad news - no point in moving further */ - } - struct neuron_device *nd = neuron_pci_get_device(private_data->device_idx); if (nd == NULL) { pr_err("ndmabuf_unmap: Failed to retrieve nd%d, is the device closed?\n", @@ -280,28 +197,36 @@ static void ndmabuf_release(struct dma_buf *dmabuf) static const struct dma_buf_ops ndmabuf_ops = { .attach = ndmabuf_attach, - .detach = ndmabuf_detach, .map_dma_buf = ndmabuf_map, .unmap_dma_buf = ndmabuf_unmap, .release = ndmabuf_release, }; /* Create a new dmabuf object and retrieve its fd */ -int ndmabuf_get_fd(u64 va, u64 size, int *dmabuf_fd) +int ndmabuf_get_fd(u64 va, u64 size, int *dmabuf_fd, u64 *offset) { struct dma_buf *dmabuf; struct ndmabuf_private_data *private_data; + struct nmmap_node *mmap; int device_idx; bool mmap_found; int fd, ret; + u64 va_end, aligned_va, aligned_size, page_offset; private_data = kzalloc(sizeof(struct ndmabuf_private_data), GFP_KERNEL); if (private_data == NULL) { return -ENOMEM; } - private_data->is_attached = 0; private_data->va = (void *)va; - private_data->size = size; + + aligned_va = ALIGN_DOWN(va, PAGE_SIZE); + page_offset = va - aligned_va; + if (check_add_overflow(page_offset, size, &aligned_size)) { + ret = -EINVAL; + goto err_free_private_data; + } + + private_data->aligned_va = (void *)aligned_va; /* Detect invalid VA/size by iterating over all available neuron devices to * find the matching mmap node */ @@ -312,10 +237,17 @@ int ndmabuf_get_fd(u64 va, u64 size, int *dmabuf_fd) continue; write_lock(&nd->mpset.rbmmaplock); - struct nmmap_node *mmap = nmmap_search_va(nd, private_data->va); + mmap = nmmap_search_va(nd, private_data->va); if (mmap != NULL) { + if (check_add_overflow(va, size, &va_end) || ((va_end - (u64)mmap->va) > mmap->size)) { + write_unlock(&nd->mpset.rbmmaplock); + pr_err("ndmabuf_get_fd: va + size exceeds mmap region\n"); + ret = -EINVAL; + goto err_free_private_data; + } + private_data->device_idx = device_idx; + mmap_found = true; write_unlock(&nd->mpset.rbmmaplock); - mmap_found = 1; break; } write_unlock(&nd->mpset.rbmmaplock); @@ -328,14 +260,14 @@ int ndmabuf_get_fd(u64 va, u64 size, int *dmabuf_fd) goto err_free_private_data; } - private_data->device_idx = device_idx; - DEFINE_DMA_BUF_EXPORT_INFO(exp_info); exp_info.ops = &ndmabuf_ops; - exp_info.size = size; + exp_info.size = aligned_size; exp_info.flags = O_CLOEXEC; exp_info.priv = private_data; + /* On success, dma_buf_export assigns ownership of private_data to the + * dmabuf object via exp_info.priv; it will be freed in ndmabuf_release. */ dmabuf = dma_buf_export(&exp_info); if (IS_ERR(dmabuf)) { ret = PTR_ERR(dmabuf); @@ -347,21 +279,24 @@ int ndmabuf_get_fd(u64 va, u64 size, int *dmabuf_fd) if (fd < 0) { if (fd == -EMFILE) { pr_err("dma_buf_fd failed: too many open files\n"); - ret = -EMFILE; } else { - pr_err("error %d while installing a file descriptor for dma-buf\n", ret); - ret = -EINVAL; + pr_err("error %d while installing a file descriptor for dma-buf\n", fd); } + ret = fd; goto err_dma_buf_put; } - private_data->fd = fd; *dmabuf_fd = fd; + *offset = page_offset; return 0; err_dma_buf_put: + /* dma_buf_put drops the last file ref, triggering dma_buf_release -> + * ndmabuf_release -> kfree(private_data). Must not fall through to + * err_free_private_data or private_data will be freed twice. */ dma_buf_put(dmabuf); + return ret; err_free_private_data: kfree(private_data); @@ -370,7 +305,7 @@ int ndmabuf_get_fd(u64 va, u64 size, int *dmabuf_fd) #else // #if LINUX_VERSION_CODE >= KERNEL_VERSION(5, 15, 0) -int ndmabuf_get_fd(u64 va, u64 size, int *dmabuf_fd) +int ndmabuf_get_fd(u64 va, u64 size, int *dmabuf_fd, u64 *offset) { return -EPROTONOSUPPORT; } diff --git a/neuron_dmabuf.h b/neuron_dmabuf.h index d1887ef..b369a3e 100644 --- a/neuron_dmabuf.h +++ b/neuron_dmabuf.h @@ -19,9 +19,10 @@ * @addr: Device buffer virtual address * @size: Device buffer size (in bytes) * @fd: dma-buf fd + * @offset: offset of virtual address within dmabuf * * Return: 0 on success, error code on failure */ -int ndmabuf_get_fd(u64 va, u64 size, int *fd); +int ndmabuf_get_fd(u64 va, u64 size, int *fd, u64 *offset); #endif diff --git a/neuron_ds.c b/neuron_ds.c index b906c0c..fc63e18 100644 --- a/neuron_ds.c +++ b/neuron_ds.c @@ -21,7 +21,7 @@ int neuron_ds_init(struct neuron_datastore *nds, struct neuron_device *parent) mutex_init(&nds->lock); for (idx = 0; idx < NEURON_MAX_DATASTORE_ENTRIES_PER_DEVICE; idx++) { ret = mc_alloc_align(parent, MC_LIFESPAN_DEVICE, NEURON_DATASTORE_SIZE, 0, - MEM_LOC_HOST, 0, 0, 0, NEURON_MEMALLOC_TYPE_NCDEV_HOST, + MEM_LOC_HOST, 0, 0, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &nds->entries[idx].mc); if (ret) { pr_err("nds allocation failure for nd[%d]", parent->device_index); diff --git a/neuron_fw_io.c b/neuron_fw_io.c index 186c390..4bf58af 100644 --- a/neuron_fw_io.c +++ b/neuron_fw_io.c @@ -820,19 +820,19 @@ static inline uint32_t repairable_ecc_err_count(uint32_t api_version, uint32_t e void fw_io_get_total_ecc_err_counts(void *bar0, uint32_t *unrepairable_ecc_count, uint32_t *repairable_ecc_count) { uint32_t total_unrepairable_ecc_err_count = 0; uint32_t total_repairable_ecc_err_count = 0; - uint32_t channel = 0; + uint32_t hbm_index = 0; uint32_t ecc_err_count = 0; uint64_t ecc_offset = 0; uint32_t api_version; fw_io_api_version_read(bar0, &api_version); - for (channel = 0; channel < ndhal->ndhal_address_map.dram_channels; channel++) { - ecc_offset = FW_IO_REG_HBM0_ECC_OFFSET + channel * sizeof(uint32_t); + for (hbm_index = 0; hbm_index < ndhal->ndhal_address_map.num_hbms; hbm_index++) { + ecc_offset = FW_IO_REG_HBM0_ECC_OFFSET + hbm_index * sizeof(uint32_t); ecc_err_count = 0; int ret = fw_io_ecc_read(bar0, ecc_offset, &ecc_err_count); if (ret) { - pr_err("sysfs failed to read ECC HBM%u error from FWIO\n", channel); + pr_err("sysfs failed to read ECC HBM%u error from FWIO\n", hbm_index); } else if (ecc_err_count != 0xdeadbeef) { total_unrepairable_ecc_err_count += unrepairable_ecc_err_count(api_version, ecc_err_count); total_repairable_ecc_err_count += repairable_ecc_err_count(api_version, ecc_err_count); diff --git a/neuron_ioctl.h b/neuron_ioctl.h index fc0c342..095dbb4 100644 --- a/neuron_ioctl.h +++ b/neuron_ioctl.h @@ -16,8 +16,8 @@ struct neuron_ioctl_mem_alloc { __u64 size; // [in] Allocation size __u32 host_memory; // [in] If true allocates from host memory; else allocates from device memory - __u32 dram_channel; // [in] DRAM channel in device memory - __u32 dram_region; // [in] DRAM region in device memory + __u32 hbm_index; // [in] HBM index within a device + __u32 dram_region_unused; __u32 nc_id; // [in] NeuronCore id(valid only if location is device) __u64 *mem_handle; // [out] Allocated memory handle would stored here. }; @@ -26,8 +26,8 @@ struct neuron_ioctl_mem_alloc_v2 { __u64 size; // [in] Allocation size __u64 align; // [in] alignment __u32 host_memory; // [in] If true allocates from host memory; else allocates from device memory - __u32 dram_channel; // [in] DRAM channel in device memory - __u32 dram_region; // [in] DRAM region in device memory + __u32 hbm_index; // [in] HBM index within a device + __u32 dram_region_unused; __u32 nc_id; // [in] NeuronCore id(valid only if location is device) __u64 *mem_handle; // [out] Allocated memory handle would stored here. }; @@ -36,8 +36,8 @@ struct neuron_ioctl_mem_alloc_v2_mem_type { __u64 size; // [in] Allocation size __u64 align; // [in] alignment __u32 host_memory; // [in] If true allocates from host memory; else allocates from device memory - __u32 dram_channel; // [in] DRAM channel in device memory - __u32 dram_region; // [in] DRAM region in device memory + __u32 hbm_index; // [in] HBM index within a device + __u32 dram_region_unused; __u32 nc_id; // [in] NeuronCore id(valid only if location is device) __u32 mem_type; // [in] type of allocation __u64 *mem_handle; // [out] Allocated memory handle would stored here. @@ -53,8 +53,8 @@ struct neuron_ioctl_mem_alloc_v2_mem_type64 { __u64 size; // [in] Allocation size __u64 align; // [in] alignment __u32 host_memory; // [in] If true allocates from host memory; else allocates from device memory - __u32 dram_channel; // [in] DRAM channel in device memory - __u32 dram_region; // [in] DRAM region in device memory + __u32 hbm_index; // [in] HBM index within a device + __u32 dram_region_unused; __u32 nc_id; // [in] NeuronCore id(valid only if location is device) __u32 mem_type; // [in] type of allocation __u64 *mem_handle; // [out] Allocated memory handle would stored here. @@ -70,8 +70,8 @@ struct neuron_ioctl_mem_alloc_v2_mem_type64_pa { __u64 size; // [in] Allocation size __u64 align; // [in] alignment __u32 host_memory; // [in] If true allocates from host memory; else allocates from device memory - __u32 dram_channel; // [in] DRAM channel in device memory - __u32 dram_region; // [in] DRAM region in device memory + __u32 hbm_index; // [in] HBM index within a device + __u32 dram_region_unused; __u32 nc_id; // [in] NeuronCore id(valid only if location is device) __u32 mem_type; // [in] type of allocation __u64 *mem_handle; // [out] Allocated memory handle would stored here. @@ -212,7 +212,21 @@ struct neuron_ioctl_mem_buf_copy64zc { __u32 is_copy_to_device; // [in] if set to True copies to device __u32 bar4_wr_threshold; // [in] threshold below which we will use bar4 direct write vs. DMA. Subject to driver limits. __s32 h2t_qid; // [in] h2t queue to use for the transfer. -1 = use default - __u32 dummy; // [na] pad to change size of struct to version ioctl + __u32 dummy; // [in] Explicit padding for 64-bit alignment. + __u64 sequence_num; // [in] The sequence number that uniquely identifies each async I/O. + void *context; // [in] Opaque context pointer passed back and representing a pointer to xu_error_list_t in runtime. +}; + +// Old drivers use the same zerocopy ioctl number with this smaller payload. +struct neuron_ioctl_mem_buf_copy64zc_deprecated { + __u64 mem_handle; // [in] Source or Destination memory handle from/to data needs to be copied. + void *buffer; // [in] Buffer from/to where data to be copied. + __u64 size; // [in] Size of the data to be copied. + __u64 offset; // [in] Offset in the memory handle where the data to be written/read. + __u32 is_copy_to_device; // [in] if set to True copies to device + __u32 bar4_wr_threshold; // [in] threshold below which we will use bar4 direct write vs. DMA. Subject to driver limits. + __s32 h2t_qid; // [in] h2t queue to use for the transfer. -1 = use default + __u32 dummy; // [in] Explicit padding for 64-bit alignment. }; struct neuron_ioctl_mem_buf_copy64zc_batches { @@ -375,8 +389,8 @@ struct neuron_ioctl_notifications_init_v2 { __u32 engine_index; // [in] Engine Index. __u32 size; // [in] Notification queue size in bytes __u32 on_host_memory; // [in] If true allocates NQ in host memory; else allocates in device memory - __u32 dram_channel; // [in] DRAM channel in device memory - __u32 dram_region; // [in] DRAM region in device memory + __u32 hbm_index; // [in] HBM index within a device + __u32 dram_region_unused; __u64 mmap_offset; // [out] mmap() offset for this NQ __u64 mem_handle; // [out] mem_handle for this NQ }; @@ -388,8 +402,8 @@ struct neuron_ioctl_notifications_init_with_realloc_v2 { __u32 engine_index; // [in] Engine Index. __u32 size; // [in] Notification queue size in bytes __u32 on_host_memory; // [in] If true allocates NQ in host memory; else allocates in device memory - __u32 dram_channel; // [in] DRAM channel in device memory - __u32 dram_region; // [in] DRAM region in device memory + __u32 hbm_index; // [in] HBM index within a device + __u32 dram_region_unused; __u32 force_alloc_mem; // If true force allocates new memory (and deletes already allocated memory, if any) __u64 mmap_offset; // [out] mmap() offset for this NQ __u64 mem_handle; // [out] mem_handle for this NQ @@ -509,6 +523,13 @@ struct neuron_ioctl_dmabuf_fd { __s32 *fd; }; +struct neuron_ioctl_dmabuf_fd_v2 { + __u64 va; + __u64 size; + __s32 *fd; + __u64 *offset; +}; + #define NEURON_DEVICE_DRIVER_INFO_VERSION0 0 struct neuron_ioctl_device_driver_info{ @@ -882,6 +903,7 @@ struct neuron_ioctl_host_mem_unpin { #define NEURON_IOCTL_POD_CTRL _IOWR(NEURON_IOCTL_BASE, 123, struct neuron_ioctl_pod_ctrl) #define NEURON_IOCTL_POD_CTRL_V2 _IOWR(NEURON_IOCTL_BASE, 123, struct neuron_ioctl_pod_ctrl_v2) +#define NEURON_IOCTL_MEM_BUF_ZEROCOPY64_DEPRECATED _IOWR(NEURON_IOCTL_BASE, 124, struct neuron_ioctl_mem_buf_copy64zc_deprecated) #define NEURON_IOCTL_MEM_BUF_ZEROCOPY64 _IOWR(NEURON_IOCTL_BASE, 124, struct neuron_ioctl_mem_buf_copy64zc) #define NEURON_IOCTL_H2T_DMA_ALLOC_QUEUES _IOWR(NEURON_IOCTL_BASE, 125, struct neuron_ioctl_h2t_dma_alloc_queues) @@ -910,4 +932,7 @@ struct neuron_ioctl_host_mem_unpin { #define NEURON_IOCTL_HOST_MEM_PIN _IOWR(NEURON_IOCTL_BASE, 136, struct neuron_ioctl_host_mem_pin) #define NEURON_IOCTL_HOST_MEM_UNPIN _IOW(NEURON_IOCTL_BASE, 137, struct neuron_ioctl_host_mem_unpin) +/** Get dma-buf file-descriptor with offset (v2) */ +#define NEURON_IOCTL_DMABUF_FD_V2 _IOR(NEURON_IOCTL_BASE, 138, struct neuron_ioctl_dmabuf_fd_v2 *) + #endif diff --git a/neuron_mc_handle.c b/neuron_mc_handle.c index 3b49ae9..9ffbfba 100644 --- a/neuron_mc_handle.c +++ b/neuron_mc_handle.c @@ -129,6 +129,15 @@ int nmch_handle_alloc(struct neuron_device *nd, struct mem_chunk *mc, neuron_mc_ mutex_lock(&nd->nmch.lock); + // if a handle has already been allocated for this mc (possibly by a + // concurrent caller that won the race), return it instead of allocating + // a second handle that would never be freed. + if (mc->mc_handle != NMCH_INVALID_HANDLE) { + *mc_handle = mc->mc_handle; + ret = 0; + goto done; + } + if (nmch_service_is_down(nd)) { ret = -ENOENT; goto done; diff --git a/neuron_mempool.c b/neuron_mempool.c index b5b2d2f..f296fe7 100644 --- a/neuron_mempool.c +++ b/neuron_mempool.c @@ -127,14 +127,13 @@ static void mc_remove_node(struct rb_root *root, struct mem_chunk *mc) * @start_addr: starting address of the pool * @pool_size: size of the pool. * @mem_location: location of the backing memory. - * @dram_channel: device dram channel backing this pool. - * @dram_region: device dram region backing this pool. + * @hbm_index: device hbm index backing this pool. * * Return: 0 if pool is created, a negative error code otherwise. */ static int mp_init_device_mem(struct neuron_mempool *mp, struct neuron_mempool_set *mpset, - u64 start_addr, size_t pool_size, u32 dram_channel, u32 dram_region) + u64 start_addr, size_t pool_size, u32 hbm_index) { int ret; ulong small_pool_size = mempool_small_pool_size; @@ -162,8 +161,7 @@ static int mp_init_device_mem(struct neuron_mempool *mp, struct neuron_mempool_s mp->mpset = mpset; mp->mem_location = MEM_LOC_DEVICE; - mp->dram_channel = dram_channel; - mp->dram_region = dram_region; + mp->hbm_index = hbm_index; mp->main_pool_end_addr = start_addr + main_pool_size; mp->small_pool_size = small_pool_size; @@ -194,7 +192,7 @@ static int mp_init_device_mem(struct neuron_mempool *mp, struct neuron_mempool_s mp->gen_pool_small = NULL; } - snprintf(mp->name, sizeof(mp->name), "device mempool [%d:%d]", dram_channel, dram_region); + snprintf(mp->name, sizeof(mp->name), "device mempool [%d]", hbm_index); mp->region_size = pool_size; mp->initialized = 1; @@ -329,8 +327,7 @@ static void mp_destroy_gen_pool(struct neuron_mempool *mp) static int mpset_block_carveout_regions(struct neuron_device *nd, struct neuron_mempool_set *mpset, u64 *device_dram_addr, u64 *device_dram_size) { int ret; - u64 region_sz; - int channel = 0, region = 0; + int hbm_index = 0; /* * Block carve out regions: Upper 16 MB is used internally by firmware @@ -344,24 +341,21 @@ static int mpset_block_carveout_regions(struct neuron_device *nd, struct neuron_ * * lib/genalloc.c: fix allocation of aligned buffer from non-aligned chunk */ - for (channel = 0; channel < mpset->num_channels; channel++) { - region_sz = device_dram_size[channel] / mpset->mp_device_num_regions; - for (region = 0; region < mpset->mp_device_num_regions; region++) { - const dma_addr_t start_addr = device_dram_addr[channel] + (region * region_sz); - struct mem_chunk *mc = NULL; - u32 nc_id = channel; - ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, MEMPOOL_CARVEOUT_SIZE, 0, MEM_LOC_DEVICE, channel, region, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_DEVICE, &mc); - if (ret) { - pr_err("failed to allocate hbm carveout region: ret=%d\n", ret); - return -ENOMEM; - } - if (mc->pa != start_addr) { - pr_err("carve out mc not offset 0!"); - mc_free(&mc); - return -EINVAL; - } + for (hbm_index = 0; hbm_index < mpset->num_hbms; hbm_index++) { + const dma_addr_t start_addr = device_dram_addr[hbm_index]; + struct mem_chunk *mc = NULL; + u32 nc_id = hbm_index; // TODO is this right? + ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, MEMPOOL_CARVEOUT_SIZE, 0, MEM_LOC_DEVICE, hbm_index, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_DEVICE, &mc); + if (ret) { + pr_err("failed to allocate hbm carveout: ret=%d\n", ret); + return -ENOMEM; + } + if (mc->pa != start_addr) { + pr_err("carve out mc not offset 0!"); + mc_free(&mc); + return -EINVAL; } - ndhal->ndhal_mpset.device_dram_effective_base_addr[channel] = device_dram_addr[channel] + MEMPOOL_CARVEOUT_SIZE; + ndhal->ndhal_mpset.device_dram_effective_base_addr[hbm_index] = device_dram_addr[hbm_index] + MEMPOOL_CARVEOUT_SIZE; } return 0; @@ -378,22 +372,19 @@ static int mpset_block_carveout_regions(struct neuron_device *nd, struct neuron_ static int mpset_init_device_pools(struct neuron_mempool_set *mpset, struct neuron_device *nd) { int ret; - int channel = 0, region = 0; - u64 region_sz = 0; - u64 device_dram_addr[MAX_DRAM_CHANNELS]; - u64 device_dram_size[MAX_DRAM_CHANNELS]; + int hbm_index = 0; + u64 device_dram_addr[MAX_NUM_HBMS]; + u64 device_dram_size[MAX_NUM_HBMS]; ndhal->ndhal_mpset.mpset_set_dram_and_mpset_info(mpset, device_dram_addr, device_dram_size); - for (channel = 0; channel < mpset->num_channels; channel++) { - region_sz = device_dram_size[channel] / mpset->mp_device_num_regions; - for (region = 0; region < mpset->mp_device_num_regions; region++) { - dma_addr_t addr = device_dram_addr[channel] + (region * region_sz); - ret = mp_init_device_mem(&mpset->mp_device[channel][region], mpset, addr, region_sz, channel, region); - if (ret) { - pr_err("mpset device init failed %d\n", ret); - goto fail; - } + for (hbm_index = 0; hbm_index < mpset->num_hbms; hbm_index++) { + dma_addr_t addr = device_dram_addr[hbm_index]; + u64 sz = device_dram_size[hbm_index]; + ret = mp_init_device_mem(&mpset->mp_device[hbm_index], mpset, addr, sz, hbm_index); + if (ret) { + pr_err("mpset device init failed %d\n", ret); + goto fail; } } @@ -405,10 +396,8 @@ static int mpset_init_device_pools(struct neuron_mempool_set *mpset, struct neur return 0; fail: - for (; channel >= 0; channel--) { - for (; region >= 0; region--) { - mp_destroy_gen_pool(&mpset->mp_device[channel][region]); - } + for (hbm_index--; hbm_index >= 0; hbm_index--) { + mp_destroy_gen_pool(&mpset->mp_device[hbm_index]); } memset(mpset, 0, sizeof(struct neuron_mempool_set)); @@ -498,7 +487,7 @@ static struct list_head * mpset_get_lifespan_head(struct neuron_mempool_set *mps void mpset_destructor(struct neuron_mempool_set *mpset) { - int i, channel, region; + int i, hbm_index; struct list_head *head; struct neuron_device *nd; @@ -521,10 +510,8 @@ void mpset_destructor(struct neuron_mempool_set *mpset) mp_destroy_gen_pool(&mpset->mp_hrm[i]); } - for (channel = 0; channel < mpset->num_channels; channel++) { - for (region = 0; region < mpset->mp_device_num_regions; region++) { - mp_destroy_gen_pool(&mpset->mp_device[channel][region]); - } + for (hbm_index = 0; hbm_index < mpset->num_hbms; hbm_index++) { + mp_destroy_gen_pool(&mpset->mp_device[hbm_index]); } mutex_unlock(&mpset->lock); } @@ -629,7 +616,7 @@ static inline u64 get_offset_for_scratchpad_alloc(const struct neuron_mempool *m } static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan, u64 size, u64 align, - enum mem_location location, u32 channel, u32 region, u32 nc_id, mem_alloc_category_t mem_type, + enum mem_location location, u32 hbm_index, u32 nc_id, mem_alloc_category_t mem_type, struct mem_chunk **result) { struct mem_chunk *mc; @@ -648,7 +635,7 @@ static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan return -EINVAL; } size = roundup(size, PAGE_SIZE); - if (channel >= ndhal->ndhal_address_map.dram_channels) + if (hbm_index >= ndhal->ndhal_address_map.num_hbms) return -EINVAL; #ifdef CONFIG_FAULT_INJECTION if (should_fail(&neuron_fail_mc_alloc, 1)) @@ -658,13 +645,6 @@ static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan return -EINVAL; } - if (mpset->mp_device_num_regions == 1) // shared DRAM mode, always use region 0 - region = 0; - - if (region >= mpset->mp_device_num_regions) { - return -EINVAL; - } - mc = (struct mem_chunk *)kmalloc(sizeof(struct mem_chunk), GFP_KERNEL); if (mc == NULL) return -ENOMEM; @@ -708,7 +688,7 @@ static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan else pr_info("host mem occupied %lld\n", mpset->host_mem_size); } else { - mp = &mpset->mp_device[channel][region]; + mp = &mpset->mp_device[hbm_index]; if (!mp->gen_pool) { pr_err("mempool not initialized\n"); ret = -ENOMEM; @@ -730,7 +710,7 @@ static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan mc->va = (void *)gen_pool_alloc_algo(pool, size, gen_pool_fixed_alloc, &offset_data); if (mc->va == NULL) { - pr_err("nd %d HBM %d: Could not allocate %lld bytes at offset %ld for contiguous scratchpad\n", nd->device_index, channel, size, offset_data.offset); + pr_err("nd %d HBM %d: Could not allocate %lld bytes at offset %ld for contiguous scratchpad\n", nd->device_index, hbm_index, size, offset_data.offset); ret = -ENOMEM; goto exit; } @@ -750,7 +730,7 @@ static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan if (mc->va == NULL && alt_pool != NULL) { bool pool_is_main = (pool == mp->gen_pool); pr_warn_once("no space in %s genpool for HBM %d, needed %lld (alignment %lld), trying %s genpool\n", pool_is_main ? "main":"small", - channel, size, align, pool_is_main ? "small":"main"); + hbm_index, size, align, pool_is_main ? "small":"main"); mc->va = (void *)gen_pool_alloc_algo(alt_pool, size, gen_pool_first_fit_align, &align_data); mc->gen_pool = alt_pool; @@ -771,7 +751,7 @@ static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan if (mc->va == NULL && alt_pool != NULL) { bool pool_is_main = (pool == mp->gen_pool); pr_warn_once("no space in %s genpool for HBM %d, needed %lld, trying %s genpool\n", pool_is_main ? "main":"small", - channel, size, pool_is_main ? "small":"main"); + hbm_index, size, pool_is_main ? "small":"main"); mc->va = gen_pool_dma_alloc(alt_pool, size, &mc->pa); mc->gen_pool = alt_pool; } else { @@ -788,8 +768,6 @@ static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan pr_info("%s total %ld occupied %ld needed %lld available - main gen pool %ld, small gen pool %ld\n", mp->name, mp->region_size, mp->allocated_size, size, gen_pool_avail(mp->gen_pool), gen_pool_avail(mp->gen_pool_small)); } - pr_info("device regions %d occupied %lld\n", mpset->mp_device_num_regions, - mpset->device_mem_size); } } if (mc->va == NULL) { @@ -803,8 +781,7 @@ static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan mc->size = size; mc->mc_handle = NMCH_INVALID_HANDLE; mc->mem_location = location; - mc->dram_channel = channel; - mc->dram_region = region; + mc->hbm_index = hbm_index; mc->nc_id = nc_id; mc->pid = task_tgid_nr(current); mc->ref_count = 1; @@ -844,9 +821,9 @@ static int mc_alloc_internal(struct neuron_device *nd, enum mc_lifespan lifespan } int mc_alloc_align(struct neuron_device *nd, enum mc_lifespan lifespan, u64 size, u64 align, - enum mem_location location, u32 channel, u32 region, u32 nc_id, mem_alloc_category_t mem_type, + enum mem_location location, u32 hbm_index, u32 nc_id, mem_alloc_category_t mem_type, struct mem_chunk **result) { - return mc_alloc_internal(nd, lifespan, size, align, location, channel, region, nc_id, mem_type, result); + return mc_alloc_internal(nd, lifespan, size, align, location, hbm_index, nc_id, mem_type, result); }; void mc_inc_refcount(struct mem_chunk *mc) @@ -898,7 +875,7 @@ void mc_free(struct mem_chunk **mcp) nsysfsmetric_dec_counter(mpset->nd, NON_NDS_METRIC, NON_NDS_COUNTER_HOST_MEM, mc->nc_id, mc->size, false); } else if (mc->mem_location == MEM_LOC_DEVICE) { struct neuron_mempool *mp; - mp = &mpset->mp_device[mc->dram_channel][mc->dram_region]; + mp = &mpset->mp_device[mc->hbm_index]; gen_pool_free(mc->gen_pool, (u64)mc->va, mc->size); mp->allocated_size -= mc->size; mpset->device_mem_size -= mc->size; @@ -928,7 +905,7 @@ void mc_free(struct mem_chunk **mcp) kfree(mc); } -int mc_dump_all_chunks(struct neuron_device *nd, u32 channel, u32 num_entries_in, struct neuron_ioctl_mem_chunk_info *data, u32 *num_entries_out) +int mc_dump_all_chunks(struct neuron_device *nd, u32 hbm_index, u32 num_entries_in, struct neuron_ioctl_mem_chunk_info *data, u32 *num_entries_out) { struct neuron_mempool_set *mpset = &nd->mpset; u32 cnt = 0; @@ -937,7 +914,7 @@ int mc_dump_all_chunks(struct neuron_device *nd, u32 channel, u32 num_entries_in read_lock(&mpset->rblock); for (node = rb_first(&mpset->root); node; node = rb_next(node)) { struct mem_chunk *mc = rb_entry(node, struct mem_chunk, node); - if (mc->mem_location == MEM_LOC_DEVICE && mc->dram_channel == channel) { + if (mc->mem_location == MEM_LOC_DEVICE && mc->hbm_index == hbm_index) { if (cnt < num_entries_in) { struct neuron_ioctl_mem_chunk_info *i = &data[cnt]; i->pa = mc->pa; @@ -950,4 +927,4 @@ int mc_dump_all_chunks(struct neuron_device *nd, u32 channel, u32 num_entries_in read_unlock(&mpset->rblock); *num_entries_out = cnt; return 0; -} \ No newline at end of file +} diff --git a/neuron_mempool.h b/neuron_mempool.h index 61cdd5f..44be8fd 100644 --- a/neuron_mempool.h +++ b/neuron_mempool.h @@ -46,8 +46,7 @@ struct neuron_mempool { struct neuron_mempool_set *mpset; // parent mpset enum mem_location mem_location; // location of the memory - u32 dram_channel; // DRAM channel valid only if location is device - u32 dram_region; // DRAM region valid only if location is device + u32 hbm_index; // HBM index valid only if location is device struct gen_pool *gen_pool; // main gen pool struct gen_pool *gen_pool_small; // small gen pool for small allocations to avoid fragmentation, may be NULL @@ -66,9 +65,7 @@ struct neuron_mempool { u64 scratchpad_size; // only used for allocations of type NEURON_MEMALLOC_TYPE_CONTIGUOUS_SCRATCHPAD_DEVICE }; -// DRAM region is split into multiple regions. -#define MAX_DDR_REGIONS 4 -#define MAX_DRAM_CHANNELS 4 +#define MAX_NUM_HBMS 4 // max supported number of HBMs per device, VX_NUM_HBMS macros define it per specific chip // start page size for host MP #define MP_HOST_PAGE_SIZE_MIN (256UL * 1024) @@ -80,9 +77,8 @@ struct neuron_mempool_set { struct neuron_device *nd; // backponter to neuron_device - u32 mp_device_num_regions; // number of regions in the device pool - u32 num_channels; // number of regions in the device pool - struct neuron_mempool mp_device[MAX_DRAM_CHANNELS][MAX_DDR_REGIONS]; // device memory pools + u32 num_hbms; // number of HBMs in the device pool + struct neuron_mempool mp_device[MAX_NUM_HBMS]; // device memory pools struct neuron_mempool mp_hrm[MP_HOST_RESERVE_MEMORY_POOL_COUNT]; // host reserve memory pools @@ -131,8 +127,7 @@ struct mem_chunk { struct neuron_mempool_set *mpset; // back pointer to mpset struct gen_pool *gen_pool; // pointer to genpool - u32 dram_channel; // DRAM channel - u32 dram_region; // TDRAM region + u32 hbm_index; // HBM index within the device u32 nc_id; //neuron core index neuron_mc_handle_t mc_handle; // memchunk handle mem_alloc_category_t alloc_type; // memory allocation category @@ -186,15 +181,14 @@ struct mem_chunk *mpset_search_mc(struct neuron_mempool_set *mp, phys_addr_t pa) * @size: Allocation size * @align: alignment requirement * @location: Backing DRAM location(host/device) - * @channel: Backing DRAM channel - * @region: Region in the backing DRAM + * @hbm_index: Backing HBM * @mem_type: category of memory allocation (used for sysfs memory counters) * @result: Buffer to store the allocated memory chunk pointer * * Return: 0 if allocation succeeds, a negative error code otherwise. */ int mc_alloc_align(struct neuron_device *nd, enum mc_lifespan lifespan, u64 size, u64 align, - enum mem_location location, u32 channel, u32 region, u32 nc_id, mem_alloc_category_t mem_type, + enum mem_location location, u32 hbm_index, u32 nc_id, mem_alloc_category_t mem_type, struct mem_chunk **result); /** @@ -219,8 +213,7 @@ void mpset_free_expired_mc(struct neuron_mempool_set *mpset, enum mc_lifespan li */ void mc_inc_refcount(struct mem_chunk *mc); -//int mc_dump_all_chunks(struct neuron_device *nd, u32 channel); -int mc_dump_all_chunks(struct neuron_device *nd, u32 channel, u32 num_entries_in, struct neuron_ioctl_mem_chunk_info *data, u32 *num_entries_out); +int mc_dump_all_chunks(struct neuron_device *nd, u32 hbm_index, u32 num_entries_in, struct neuron_ioctl_mem_chunk_info *data, u32 *num_entries_out); static inline bool mc_access_is_within_bounds(const struct mem_chunk *mc, u64 access_offset, u64 access_size) { diff --git a/neuron_mmap.c b/neuron_mmap.c index a4faf63..0106332 100644 --- a/neuron_mmap.c +++ b/neuron_mmap.c @@ -502,7 +502,7 @@ int nmmap_get_va_placement(void *va, int *device_index, int *hbm_index) } write_unlock(&nd->mpset.rbmmaplock); if (hbm_pa != (u64)-1) { // found it, now find HBM - for (hbm = 0; hbm < ndhal->ndhal_address_map.dram_channels; hbm++) { + for (hbm = 0; hbm < ndhal->ndhal_address_map.num_hbms; hbm++) { u64 start = ndhal->ndhal_mpset.device_dram_effective_base_addr[hbm]; u64 end = ndhal->ndhal_mpset.device_dram_end_addr[hbm]; if (hbm_pa >= start && hbm_pa < end) { diff --git a/neuron_module.c b/neuron_module.c index f397aa9..48feed0 100644 --- a/neuron_module.c +++ b/neuron_module.c @@ -20,13 +20,13 @@ #include "neuron_dma.h" #include "neuron_test.h" -MODULE_DESCRIPTION("Neuron Driver, built from SHA: 38e99b1fb8df603ee4109868c6d949e30f2c32c8"); +MODULE_DESCRIPTION("Neuron Driver, built from SHA: c64fcb38034831fa1d30e95aca96ff5ed5005726"); MODULE_LICENSE("GPL"); -MODULE_VERSION("2.28.0.0"); +MODULE_VERSION("2.29.0.0"); MODULE_ALIAS("pci:v00001d0fd00007064sv*sd*bc*sc*i*"); -const char driver_version[] = "2.28.0.0"; -const char driver_revision[] = "38e99b1fb8df603ee4109868c6d949e30f2c32c8"; +const char driver_version[] = "2.29.0.0"; +const char driver_revision[] = "c64fcb38034831fa1d30e95aca96ff5ed5005726"; #ifdef CONFIG_FAULT_INJECTION diff --git a/neuron_nq.c b/neuron_nq.c index 676c764..932522a 100644 --- a/neuron_nq.c +++ b/neuron_nq.c @@ -35,7 +35,7 @@ static int nnq_halt(struct neuron_device *nd, u8 nc_id, u8 eng_index, u32 nq_typ return -EINVAL; nq_id = ndhal->ndhal_nq.nnq_get_nqid(nd, nc_id, eng_index, nq_type); - if (nq_id >= MAX_NQ_SUPPORTED) + if (nq_id >= ndhal->ndhal_address_map.nq_per_nc) return -EINVAL; if (nd->nq_mc[nc_id][nq_id] == NULL) { @@ -56,7 +56,7 @@ static int nnq_destroy(struct neuron_device *nd, u8 nc_id, u8 eng_index, u32 nq_ return -EINVAL; nq_id = ndhal->ndhal_nq.nnq_get_nqid(nd, nc_id, eng_index, nq_type); - if (nq_id >= MAX_NQ_SUPPORTED) + if (nq_id >= ndhal->ndhal_address_map.nq_per_nc) return -EINVAL; if (nd->nq_mc[nc_id][nq_id] == NULL) { @@ -68,9 +68,52 @@ static int nnq_destroy(struct neuron_device *nd, u8 nc_id, u8 eng_index, u32 nq_ return 0; } +void nnq_destroy_storage(struct neuron_device *nd) +{ + int i; + + if (!nd) + return; + + for (i = 0; i < MAX_NC_PER_DEVICE; i++) { + kfree(nd->nq_mc[i]); + nd->nq_mc[i] = NULL; + } + for (i = 0; i < MAX_TS_PER_DEVICE; i++) { + kfree(nd->ts_nq_mc[i]); + nd->ts_nq_mc[i] = NULL; + } +} + +int nnq_init_storage(struct neuron_device *nd) +{ + u16 sz = ndhal->ndhal_address_map.nq_per_nc; + int i; + + if (sz == 0) { + pr_err("ndhal->ndhal_address_map.nq_per_nc is 0; ndhal not initialized\n"); + return -EINVAL; + } + + for (i = 0; i < MAX_NC_PER_DEVICE; i++) { + nd->nq_mc[i] = kcalloc(sz, sizeof(struct mem_chunk *), GFP_KERNEL); + if (!nd->nq_mc[i]) + goto fail; + } + for (i = 0; i < MAX_TS_PER_DEVICE; i++) { + nd->ts_nq_mc[i] = kcalloc(sz, sizeof(struct mem_chunk *), GFP_KERNEL); + if (!nd->ts_nq_mc[i]) + goto fail; + } + return 0; +fail: + nnq_destroy_storage(nd); + return -ENOMEM; +} + int nnq_init(struct neuron_device *nd, u8 nc_id, u8 eng_index, u32 nq_type, u32 size, - u32 on_host_memory, u32 dram_channel, u32 dram_region, bool force_alloc_mem, + u32 on_host_memory, u32 hbm_index, bool force_alloc_mem, struct mem_chunk **nq_mc, u64 *mmap_offset) { // Check that size is power of 2 @@ -83,14 +126,14 @@ int nnq_init(struct neuron_device *nd, u8 nc_id, u8 eng_index, u32 nq_type, u32 return -EINVAL; u8 nq_id = ndhal->ndhal_nq.nnq_get_nqid(nd, nc_id, eng_index, nq_type); - if (nq_id >= MAX_NQ_SUPPORTED) + if (nq_id >= ndhal->ndhal_address_map.nq_per_nc) return -EINVAL; struct mem_chunk *mc = nd->nq_mc[nc_id][nq_id]; if (mc == NULL || force_alloc_mem) { struct mem_chunk *_mc = NULL; int ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, size, (on_host_memory) ? 0 : size, on_host_memory ? MEM_LOC_HOST : MEM_LOC_DEVICE, - dram_channel, dram_region, nc_id, on_host_memory ? NEURON_MEMALLOC_TYPE_NOTIFICATION_HOST : NEURON_MEMALLOC_TYPE_NOTIFICATION_DEVICE, &_mc); + hbm_index, nc_id, on_host_memory ? NEURON_MEMALLOC_TYPE_NOTIFICATION_HOST : NEURON_MEMALLOC_TYPE_NOTIFICATION_DEVICE, &_mc); if (ret) return ret; ndhal->ndhal_nq.nnq_set_hwaddr(nd, nc_id, eng_index, nq_type, size, _mc->pa); diff --git a/neuron_nq.h b/neuron_nq.h index e956a1a..6292ff3 100644 --- a/neuron_nq.h +++ b/neuron_nq.h @@ -52,6 +52,23 @@ static inline void notific_write_nq_f_size(void __iomem *base, size_t index, #define NOTIFIC_NQ_HEAD_OFFSET 0x10c +/** + * nnq_init_storage() - Allocate the per-NC and per-TS notification queue + * tracking arrays + * + * @nd: neuron device + * + */ +int nnq_init_storage(struct neuron_device *nd); + +/** + * nnq_destroy_storage() - Free the per-NC and per-TS notification queue + * tracking arrays + * + * @nd: neuron device + */ +void nnq_destroy_storage(struct neuron_device *nd); + /** * nnq_init() - Initialize notification queue for NeuronCore * @@ -61,8 +78,7 @@ static inline void notific_write_nq_f_size(void __iomem *base, size_t index, * @nq_type: type of the notification queue * @size: size of queue in bytes * @on_host_memory: if true, NQ is created in host memory - * @dram_channel: If NQ is created on device memory which DRAM channel to use. - * @dram_region: If NQ is created on device memory which DRAM region to use. + * @hbm_index: If NQ is created on device memory which DRAM channel to use. * @force_alloc_mem: If true, force allocate new memory (and delete already allocated memory, if any) * @nq_mc[out]: memchunk used by the NQ will be written here * @mc_ptr[out]: Pointer to memchunk backing this NQ @@ -70,7 +86,7 @@ static inline void notific_write_nq_f_size(void __iomem *base, size_t index, * Return: 0 on if initialization succeeds, a negative error code otherwise. */ int nnq_init(struct neuron_device *nd, u8 nc_id, u8 eng_index, u32 nq_type, u32 size, - u32 on_host_memory, u32 dram_channel, u32 dram_region, + u32 on_host_memory, u32 hbm_index, bool force_alloc_mem, struct mem_chunk **nq_mc, u64 *mmap_offset); /** @@ -90,4 +106,4 @@ void nnq_destroy_nc(struct neuron_device *nd, u8 nc_id); */ void nnq_destroy_all(struct neuron_device *nd); -#endif \ No newline at end of file +#endif diff --git a/neuron_pci.c b/neuron_pci.c index 353fa8e..b1356c6 100644 --- a/neuron_pci.c +++ b/neuron_pci.c @@ -150,6 +150,10 @@ static int neuron_pci_device_init(struct neuron_device *nd) if (ret) goto fail_mpset; + ret = nnq_init_storage(nd); + if (ret) + goto fail_nq_mc; + // Initialize CRWL struct for (i = 0; i < MAX_NC_PER_DEVICE; i++) mutex_init(&nd->crwl[i].lock); @@ -164,11 +168,15 @@ static int neuron_pci_device_init(struct neuron_device *nd) ret = nr_start_ncs(nd, NEURON_NC_MAP_DEVICE, NEURON_RESET_REQUEST_ALL); if (ret) - return ret; + goto fail_start_ncs; return 0; +fail_start_ncs: + ncdev_delete_device_node(nd); fail_chardev: + nnq_destroy_storage(nd); +fail_nq_mc: mpset_destructor(&nd->mpset); fail_mpset: nmch_handle_cleanup(nd); @@ -190,6 +198,7 @@ static int neuron_pci_device_close(struct neuron_device *nd) } // disable NQ after disabling PCI device so that the device cant DMA anything after this nnq_destroy_all(nd); + nnq_destroy_storage(nd); ndmar_close(nd); neuron_ds_destroy(&nd->datastore); mpset_destructor(&nd->mpset); @@ -445,13 +454,13 @@ static int neuron_pci_probe(struct pci_dev *dev, const struct pci_device_id *id) if (ret) goto fail_nds_resource; - ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, NDMA_QUEUE_DUMMY_RING_SIZE, 0, MEM_LOC_HOST, 0, 0, 0, NEURON_MEMALLOC_TYPE_NCDEV_HOST, + ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, NDMA_QUEUE_DUMMY_RING_SIZE, 0, MEM_LOC_HOST, 0, 0, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &nd->ndma_q_dummy_mc); if (ret) goto fail_nds_resource; // allocate memset mc (if datastore succeeded) - ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, MEMSET_HOST_BUF_SIZE, 0, MEM_LOC_HOST, 0, 0, 0, NEURON_MEMALLOC_TYPE_NCDEV_HOST, + ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, MEMSET_HOST_BUF_SIZE, 0, MEM_LOC_HOST, 0, 0, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &nd->memset_mc); if (ret) goto fail_memset_mc; diff --git a/neuron_power.c b/neuron_power.c index 58887d5..6613ec7 100644 --- a/neuron_power.c +++ b/neuron_power.c @@ -94,6 +94,51 @@ static inline u16 npower_get_sample_num(u32 power_sample) return (u16)((power_sample >> 16) & 0xFFFF); } +/** + * npower_classify_die_observation() - figure out what the per-die cache should say after + * a single power read attempt + * + * @read_ret: the return value from the firmware read (0 on success, an error code + * otherwise) + * @raw_sample: the raw power sample word from firmware. Only meaningful when @read_ret + * is 0. + * @entry: location to fill in with the cache entry we should store after this + * observation + * + * This function is pure - it doesn't touch any locks or any device state - so it's easy + * to reason about and easy to unit test. + * + * The rules: + * - If the read failed, record a READ_ERROR. + * - If the read succeeded but firmware handed us a util value that's outside the legal + * range, record a BOGUS entry holding the offending value (we keep it around so the + * sysfs row can show it for debugging). + * - Otherwise, record a GOOD entry with the new util/counter pair. + */ +static void npower_classify_die_observation(int read_ret, u32 raw_sample, + struct neuron_power_die_cache_entry *entry) +{ + u16 util; + u16 cnt; + + if (read_ret != 0) { + entry->populated = true; + entry->state = NEURON_POWER_DIE_STATE_READ_ERROR; + entry->util_bips = 0; + entry->counter = 0; + return; + } + + util = npower_get_utilization(raw_sample); + cnt = npower_get_sample_num(raw_sample); + + entry->populated = true; + entry->state = (util > NEURON_MAX_POWER_UTIL_BIPS) ? NEURON_POWER_DIE_STATE_BOGUS + : NEURON_POWER_DIE_STATE_GOOD; + entry->util_bips = util; + entry->counter = cnt; +} + /** * npower_store_utilization() - Store a sample of neuron power utilization data in the set of * samples for a device @@ -266,6 +311,41 @@ static bool npower_in_simulated_env(void) return narch_is_qemu() || narch_is_emu(); } +/** + * npower_update_die_cache() - update the per-die cache entry for a single die based on a + * freshly-read power sample from firmware + * + * @nd: pointer to the neuron device whose cache we are updating + * @die: which die's cache entry to update + * @read_ret: the return value from the firmware read for this die + * @raw_sample: the raw power sample word read from firmware. Only meaningful when + * @read_ret is 0. Low 16 bits are util_bips, high 16 bits are the sample + * counter. + * + * The caller is responsible for performing the firmware read before calling this - we + * never issue MMIO ourselves. We classify the new observation outside the lock and then + * take stats_lock just long enough to store the result, so we never hold stats_lock + * across firmware I/O and a concurrent sysfs reader can't observe a torn cache entry. + * + * Note that this overwrites the cache entry on every call. In particular, we don't + * suppress duplicate counters here the way npower_select_power() does for the aggregate + * stats - the aggregator gate exists to keep the per-minute stats from being skewed by + * stale firmware samples, but this cache just needs to mirror what the previous live + * sysfs read path used to return, including transient READ_ERROR and BOGUS states. + * Don't try to unify the two. + */ +static void __maybe_unused npower_update_die_cache(struct neuron_device *nd, unsigned die, + int read_ret, u32 raw_sample) +{ + struct neuron_power_die_cache_entry next; + + npower_classify_die_observation(read_ret, raw_sample, &next); + + npower_acquire_lock(nd); + nd->power.per_die_cache[die] = next; + npower_release_lock(nd); +} + static void npower_select_power(struct neuron_device *nd, u32 power_samples[NEURON_POWER_MAX_DIE]) { bool duplicate_read = false; @@ -328,6 +408,13 @@ int npower_sample_utilization(void *dev) for (die = 0; die < ndhal->ndhal_address_map.dice_per_device; die++) { ret[die] = fw_io_device_power_read(nd->npdev.bar0, &power_samples[die], die); + // Mirror this observation into the per-die cache that backs the raw + // sysfs read path. We do this whether or not the read succeeded, + // independently of npower_select_power's duplicate-tick gate below. + // See npower_update_die_cache for why these two paths intentionally + // don't share the same gating rule. + npower_update_die_cache(nd, die, ret[die], power_samples[die]); + if (ret[die]) { failed_read = true; // Note for later when we determine whether we have data to store @@ -425,6 +512,92 @@ int npower_format_stats(void *dev, char buffer[], unsigned bufflen) ((bytes_formatted == bufflen) && buffer[bufflen - 1] != '\0')); } +/** + * npower_emit_die_row() - format a single per-die row of the raw sysfs output from a + * cache snapshot + * + * @die: the die index to print at the start of the row + * @snapshot: pointer to a cache entry that the caller has already copied out of + * per_die_cache[] under stats_lock. We operate on the caller's snapshot, so + * we don't need to take any locks ourselves. + * @buf: where to write the row + * @bufflen: how much room is left in @buf + * + * Returns the number of bytes written, with the same scnprintf truncation semantics that + * the rest of this file uses, so the caller can just add this to its running offset. + * + * The output is one of these four forms, picked based on what the snapshot says: + * ,unavailable - we haven't gotten a sample for this die yet, or the + * snapshot is otherwise unpopulated + * ,, - normal good reading + * ,read_error - the most recent firmware read for this die failed + * ,bogus(), - firmware returned an out-of-range value; we keep the + * offending value visible to make debugging easier + */ +static unsigned __maybe_unused +npower_emit_die_row(unsigned die, + const struct neuron_power_die_cache_entry *snapshot, + char *buf, unsigned bufflen) +{ + if (!snapshot->populated) { + return scnprintf(buf, bufflen, "%u,unavailable\n", die); + } + + switch (snapshot->state) { + case NEURON_POWER_DIE_STATE_READ_ERROR: + return scnprintf(buf, bufflen, "%u,read_error\n", die); + case NEURON_POWER_DIE_STATE_BOGUS: + return scnprintf(buf, bufflen, "%u,bogus(%u),%u\n", + die, snapshot->util_bips, snapshot->counter); + case NEURON_POWER_DIE_STATE_GOOD: + default: + return scnprintf(buf, bufflen, "%u,%u,%u\n", + die, snapshot->util_bips, snapshot->counter); + } +} + +int npower_format_raw(void *dev, char buffer[], unsigned bufflen) +{ + struct neuron_device *nd = (struct neuron_device *)dev; + struct neuron_power_die_cache_entry snapshot[NEURON_POWER_MAX_DIE]; + unsigned offset = 0; + unsigned die; + unsigned dice; + + // Always emit the header first. + offset += scnprintf(buffer + offset, bufflen - offset, "die,util_bips,counter\n"); + + // Top-level unavailability: if the device isn't ready, we're in simulation, or + // firmware doesn't support power readings, just emit "unavailable" for every die. + // We don't take stats_lock or look at per_die_cache[] in that case - there's + // nothing useful in there anyway. + if (!nd || nd->device_state != NEURON_DEVICE_STATE_READY || + npower_in_simulated_env() || !npower_enabled_in_fw(nd)) { + for (die = 0; die < ndhal->ndhal_address_map.dice_per_device; die++) { + offset += scnprintf(buffer + offset, bufflen - offset, + "%u,unavailable\n", die); + } + return 0; + } + + // Snapshot the cache under stats_lock and do all the formatting outside the lock. + // Holding the lock just across a memcpy keeps it cheap, ensures no scnprintf runs + // under it, and guarantees that each formatted row reflects a single, complete + // cache write rather than a torn update. We don't issue MMIO from here at all - + // the periodic sampler is the sole producer of per-die state. + npower_acquire_lock(nd); + memcpy(snapshot, nd->power.per_die_cache, sizeof(snapshot)); + npower_release_lock(nd); + + dice = ndhal->ndhal_address_map.dice_per_device; + for (die = 0; die < dice; die++) { + offset += npower_emit_die_row(die, &snapshot[die], + buffer + offset, bufflen - offset); + } + + return 0; +} + int npower_init_stats(struct neuron_device *nd) { u16 zero_counter_array[NEURON_POWER_MAX_DIE] = { 0 }; @@ -442,6 +615,11 @@ int npower_init_stats(struct neuron_device *nd) npower_acquire_lock(nd); memset(&nd->power.current_stats, 0, sizeof(struct neuron_power_stats)); nd->power.current_stats.status = POWER_STATUS_NO_DATA; + + // Reset the per-die cache. Zeroing it leaves every entry with populated=false, + // which is the only thing the sysfs reader cares about until the sampler fills + // entries in - the other fields are just along for the ride at this point. + memset(&nd->power.per_die_cache, 0, sizeof(nd->power.per_die_cache)); npower_release_lock(nd); return 0; diff --git a/neuron_power.h b/neuron_power.h index f14ef89..daed6fe 100644 --- a/neuron_power.h +++ b/neuron_power.h @@ -81,9 +81,48 @@ struct neuron_power_stats { u16 max_power_bips; }; +/** + * Per-die state recorded by the sampler each tick. + * + * GOOD - read succeeded and util_bips <= NEURON_MAX_POWER_UTIL_BIPS. + * READ_ERROR - the firmware read returned a nonzero error. + * BOGUS - the read succeeded but util_bips was out of range. + */ +enum neuron_power_die_state { + NEURON_POWER_DIE_STATE_GOOD = 0, + NEURON_POWER_DIE_STATE_READ_ERROR, + NEURON_POWER_DIE_STATE_BOGUS, +}; + +/** + * One per-die cache entry. Written by the sampler-side cache update helper, + * read by npower_format_raw(). Access is serialized by stats_lock. + * + * @populated: False until the sampler has read power data for this die since + * the last npower_init_stats(); while false, no other field is + * meaningful and the formatter emits ",unavailable\n". + * @state: last per-die state. Only read when populated. + * @util_bips: if state == GOOD: cached in-range util reading. + * if state == BOGUS: the offending out-of-range util value (so the formatter + * can emit "bogus()"). + * if state == READ_ERROR: undefined. + * @counter: if state == GOOD or BOGUS: firmware sample counter associated with util_bips. + * if state == READ_ERROR: undefined. + */ +struct neuron_power_die_cache_entry { + bool populated; + enum neuron_power_die_state state; + u16 util_bips; + u16 counter; +}; + struct neuron_power { struct neuron_power_samples current_samples; // Unaggregated raw data for the current period struct neuron_power_stats current_stats; // Aggregated stats from the last completed period + // Per-die cache backing the raw power sysfs interface. Updated by + // the periodic sampler, read by npower_format_raw(). Serialized by + // stats_lock. + struct neuron_power_die_cache_entry per_die_cache[NEURON_POWER_MAX_DIE]; struct mutex stats_lock; }; @@ -117,6 +156,20 @@ int npower_sample_utilization(void *dev); */ int npower_format_stats(void *nd, char buffer[], unsigned int bufflen); +/** + * npower_format_raw() - Read live per-die power utilization from firmware and + * format it as CSV. Unlike npower_format_stats, this + * does not consult the per-minute aggregate; it issues a + * fresh MMIO read on every call so callers can poll at + * the firmware refresh rate. + * + * @nd: expected to point to a struct neuron_device + * @buffer / @bufflen: destination + * + * Returns 0 on success. + */ +int npower_format_raw(void *nd, char buffer[], unsigned int bufflen); + /** * npower_init_power_stats() - Initializes power stats for the specified neuron device upon boot * diff --git a/neuron_reset.c b/neuron_reset.c index c8a95b4..7f01c51 100644 --- a/neuron_reset.c +++ b/neuron_reset.c @@ -190,7 +190,11 @@ static int nr_reset_thread_fn(void *arg) if (!nd->nr.req_cmpl_head) { nd->nr.req_cmpl_head = first_request; } - ITER_COAL_REQS(request_iter, first_request, last_request, request_iter->ret = state;) + ITER_COAL_REQS(request_iter, first_request, last_request, + // Write error_code before state so nr_wait sees the error code + // once it observes the state change. smp_wmb() is needed as + // nr_wait polls req->ret without holding nr_lock. + request_iter->error_code = ret; smp_wmb(); request_iter->ret = state;) } mutex_unlock(&nd->nr.nr_lock); } @@ -203,8 +207,10 @@ int nr_create_thread(struct neuron_device *nd) init_waitqueue_head(&nd->nr.wait_queue); nd->nr.thread = kthread_run(nr_reset_thread_fn, nd, "nd%d reset", nd->device_index); if (IS_ERR_OR_NULL(nd->nr.thread)) { + int err = IS_ERR(nd->nr.thread) ? PTR_ERR(nd->nr.thread) : -ENOMEM; + nd->nr.thread = NULL; pr_err("nd%d reset thread creation failed\n", nd->device_index); - return -1; + return err; } return 0; } @@ -273,7 +279,7 @@ int nr_start_ncs(struct neuron_device *nd, uint32_t nc_map, uint32_t request_id) if (nr_find_req(nd, request_id)) { pr_err("Pending reset request for pid %u on device %u already exists!", request_id, nd->device_index); mutex_unlock(&nd->nr.nr_lock); - return 1; + return -EALREADY; } if (request_id == NEURON_RESET_REQUEST_ALL) { nd->device_state = NEURON_DEVICE_STATE_RESET; @@ -293,11 +299,14 @@ int nr_start_ncs(struct neuron_device *nd, uint32_t nc_map, uint32_t request_id) if (!req) { pr_err("Failed to allocate memory for reset request %u for nd %u", request_id, nd->device_index); mutex_unlock(&nd->nr.nr_lock); - return 1; + return -ENOMEM; } req->request_id = request_id; req->nc_map = nc_map; req->ret = NEURON_RESET_STATE_STARTED; + // No smp_wmb() needed — these writes are under nr_lock. mutex_unlock + // provides a full barrier, so nr_wait will see these values in order. + req->error_code = 0; req->next = NULL; req->prev = NULL; if (nd->nr.req_pending_tail) { @@ -335,7 +344,7 @@ int nr_wait(struct neuron_device *nd, uint32_t request_id, bool check) return 0; } else { pr_err("Invalid reset request id %u", request_id); - return 1; + return -ENOENT; } } @@ -365,8 +374,14 @@ int nr_wait(struct neuron_device *nd, uint32_t request_id, bool check) } mutex_unlock(&nd->nr.nr_lock); enum neuron_reset_state ret = req->ret; + // Pairs with smp_wmb() in nr_reset_thread_fn. Ensures error_code is + // read after ret, since we poll ret without holding nr_lock. + smp_rmb(); + int error_code = req->error_code; kfree((void *)req); - return ((ret == NEURON_RESET_STATE_COMPLETED) ? 0 : 1); + if (ret == NEURON_RESET_STATE_COMPLETED) + return 0; + return error_code ? error_code : -EIO; } bool nr_op_in_reset_wnd(uint64_t op_start_time, struct neuron_device *nd) diff --git a/neuron_reset.h b/neuron_reset.h index afac5a5..43a9025 100644 --- a/neuron_reset.h +++ b/neuron_reset.h @@ -35,6 +35,7 @@ struct neuron_reset_request { uint32_t request_id; uint32_t nc_map; volatile enum neuron_reset_state ret; + volatile int error_code; volatile struct neuron_reset_request *next; volatile struct neuron_reset_request *prev; }; @@ -59,7 +60,7 @@ struct neuron_reset { * * @nd: Neuron device which will be reset by the thread. * - * Return: 0 on success, -1 on failure + * Return: 0 on success, negative errno on failure */ int nr_create_thread(struct neuron_device *nd); @@ -84,7 +85,7 @@ void nr_start(struct neuron_device *nd); * @nc_map: Neuron core to reset (NEURON_NC_MAP_DEVICE to reset all cores) * @request_id: ID of this reset request * - * Return: 0 if reset was successfully queued, 1 otherwise. + * Return: 0 if reset was successfully queued, negative errno otherwise. */ int nr_start_ncs(struct neuron_device *nd, uint32_t nc_map, uint32_t request_id); @@ -95,7 +96,7 @@ int nr_start_ncs(struct neuron_device *nd, uint32_t nc_map, uint32_t request_id) * @request_id: The reset request id to wait for * @check: If true, return success if request_id is not in the queue. * - * Return: 0 if reset was successfully completed, 1 otherwise. + * Return: 0 if reset was successfully completed, negative errno otherwise. */ int nr_wait(struct neuron_device *nd, uint32_t request_id, bool check); diff --git a/neuron_ring.c b/neuron_ring.c index be7e868..a092acf 100644 --- a/neuron_ring.c +++ b/neuron_ring.c @@ -356,13 +356,13 @@ static int ndmar_h2t_ring_alloc(struct neuron_device *nd, int nc_id, int qid) ring->size = ring_size; ring->has_compl = false; - ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, ring_size, 0, MEM_LOC_HOST, 0, 0, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &rx_mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, ring_size, 0, MEM_LOC_HOST, 0, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &rx_mc); if (ret) { pr_err("can't allocate rx queue for H2T - size %d\n", ring_size); goto error; } - ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, ring_size, 0, MEM_LOC_HOST, 0, 0, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &tx_mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, ring_size, 0, MEM_LOC_HOST, 0, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &tx_mc); if (ret) { pr_err("can't allocate tx queue for H2T - size %d\n", ring_size); goto error; @@ -371,7 +371,7 @@ static int ndmar_h2t_ring_alloc(struct neuron_device *nd, int nc_id, int qid) ndmar_ring_set_mem_chunk(eng, qid, tx_mc, NEURON_DMA_QUEUE_TYPE_TX); ndmar_ring_set_mem_chunk(eng, qid, rx_mc, NEURON_DMA_QUEUE_TYPE_RX); - ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, sizeof(u32) * 2 * NEURON_DMA_H2T_CTX_HANDLE_CNT, 0, MEM_LOC_HOST, 0, 0, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &h2t_completion_mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, sizeof(u32) * 2 * NEURON_DMA_H2T_CTX_HANDLE_CNT, 0, MEM_LOC_HOST, 0, nc_id, NEURON_MEMALLOC_TYPE_NCDEV_HOST, &h2t_completion_mc); if (ret) { pr_err("can't allocate h2t_completion_mc memory for H2T\n"); goto error; @@ -460,7 +460,7 @@ static bool ndmar_h2t_ring_claim(struct neuron_device *nd, u32 eng_id, struct nd */ int ndmar_h2t_ring_request(struct neuron_device *nd, int nc_id, bool h2t, int *rqid) { - int ret = -1; + int ret = -EBUSY; // default if no free queue found const int eng_id = ndhal->ndhal_ndmar.ndmar_get_h2t_eng_id(nd, nc_id); struct ndma_eng *eng; struct ndma_queue *queue; @@ -469,7 +469,7 @@ int ndmar_h2t_ring_request(struct neuron_device *nd, int nc_id, bool h2t, int *r eng = ndmar_acquire_engine_nl(nd, eng_id); if (eng == NULL) - return -EINVAL; + return -ENOENT; for (qid = 0; qid < ndhal->ndhal_udma.num_queues; qid++) { if (ndhal->ndhal_ndmar.ndmar_is_h2t_def_q(nd, eng_id, qid)) diff --git a/neuron_sysfs_metrics.c b/neuron_sysfs_metrics.c index 80a4947..33381e9 100644 --- a/neuron_sysfs_metrics.c +++ b/neuron_sysfs_metrics.c @@ -185,6 +185,7 @@ static const int root_arch_node_attrs_info_tbl_cnt = sizeof(root_arch_node_attrs static const nsysfsmetric_attr_info_t power_utilization_attrs_info_tbl[] = { ATTR_INFO("utilization", NON_NDS_ID_TO_SYSFS_METRIC_ID(NON_NDS_OTHER_POWER_UTILIZATION), OTHER), + ATTR_INFO("utilization_raw", NON_NDS_ID_TO_SYSFS_METRIC_ID(NON_NDS_OTHER_POWER_UTILIZATION_RAW), OTHER), }; static const int power_utilization_attrs_info_tbl_cnt = sizeof(power_utilization_attrs_info_tbl) / sizeof(nsysfsmetric_attr_info_t); @@ -409,6 +410,15 @@ static ssize_t nsysfsmetric_show_nrt_other_metrics(struct nsysfsmetric_metrics * pr_err("sysfs failed to read power stats from FWIO, error = %d", ret); } len = nsysfsmetric_sysfs_emit(buf, "%s\n", buffer); + } else if (attr->metric_id == NON_NDS_ID_TO_SYSFS_METRIC_ID(NON_NDS_OTHER_POWER_UTILIZATION_RAW)) { + struct neuron_device *nd = container_of(sysfs_metrics, struct neuron_device, sysfs_metrics); + + char buffer[256]; + int ret = npower_format_raw(nd, buffer, sizeof(buffer)); + if (ret) { + pr_err("sysfs failed to read raw power from FWIO, error = %d", ret); + } + len = nsysfsmetric_sysfs_emit(buf, "%s", buffer); } else if (attr->metric_id == NON_NDS_ID_TO_SYSFS_METRIC_ID(NON_NDS_COUNTER_PE_ARRAY_ACTIVITY)) { struct neuron_device *nd = container_of(sysfs_metrics, struct neuron_device, sysfs_metrics); diff --git a/neuron_sysfs_metrics.h b/neuron_sysfs_metrics.h index 9c2445c..eb60681 100644 --- a/neuron_sysfs_metrics.h +++ b/neuron_sysfs_metrics.h @@ -78,6 +78,7 @@ enum nsysfsmetric_non_nds_ids { // The metrics needed by sysfs metrics but not s NON_NDS_HEALTH_STATUS_HBM_UE_COUNT, NON_NDS_HEALTH_STATUS_REPAIRABLE_HBM_UE_COUNT, NON_NDS_HEALTH_STATUS_HW_ERROR_EVENT, + NON_NDS_OTHER_POWER_UTILIZATION_RAW, }; struct neuron_device; diff --git a/neuron_topsp.c b/neuron_topsp.c index 7638fbf..21cf55f 100644 --- a/neuron_topsp.c +++ b/neuron_topsp.c @@ -47,7 +47,7 @@ int ts_nq_destroy(struct neuron_device *nd, u8 ts_id, u8 eng_index, u32 nq_type) return -EINVAL; nq_id = ndhal->ndhal_topsp.ts_nq_get_nqid(nd, eng_index, nq_type); - if (nq_id >= MAX_NQ_SUPPORTED) + if (nq_id >= ndhal->ndhal_address_map.nq_per_nc) return -EINVAL; if (nd->ts_nq_mc[ts_id][nq_id] == NULL) diff --git a/neuron_topsp.h b/neuron_topsp.h index 1c46f50..945eb26 100644 --- a/neuron_topsp.h +++ b/neuron_topsp.h @@ -16,8 +16,7 @@ * @nq_type: type of the notification queue * @size: size of queue in bytes * @on_host_memory: if true, NQ is created in host memory - * @dram_channel: If NQ is created on device memory which DRAM channel to use. - * @dram_region: If NQ is created on device memory which DRAM region to use. + * @hbm_index: If NQ is created on device memory which HBM to use. * @force_alloc_mem: If true, force allocate new memory (and delete already allocated memory, if any) * @nq_mc[out]: memchunk used by the NQ will be written here * @mc_ptr[out]: Pointer to memchunk backing this NQ @@ -25,7 +24,7 @@ * Return: 0 on if initialization succeeds, a negative error code otherwise. */ int ts_nq_init(struct neuron_device *nd, u8 ts_id, u8 eng_index, u32 nq_type, u32 size, - u32 on_host_memory, u32 dram_channel, u32 dram_region, + u32 on_host_memory, u32 hbm_index, bool force_alloc_mem, struct mem_chunk **nq_mc, u64 *mmap_offset); /** diff --git a/neuron_trace.h b/neuron_trace.h index 0782f2d..83f2a3c 100644 --- a/neuron_trace.h +++ b/neuron_trace.h @@ -187,8 +187,8 @@ TRACE_EVENT(ioctl_mem_alloc, __entry->mc->mem_location == MEM_LOC_HOST ? "HOST": "DEVICE", __entry->mc->pa, __entry->mc->size, - __entry->mc->dram_channel, - __entry->mc->dram_region, + __entry->mc->hbm_index, + 0, __entry->mc )); diff --git a/share/neuron_driver_shared.h b/share/neuron_driver_shared.h index 314f3a2..cbb73ad 100644 --- a/share/neuron_driver_shared.h +++ b/share/neuron_driver_shared.h @@ -8,6 +8,9 @@ #include #include "neuron_driver_shared_tensor_batch_op.h" +/* Feature flags are soft capabilities. Adding a feature bit does not imply a + * runtime/driver hard compatibility ID bump. + */ enum neuron_driver_feature_flag { NEURON_DRIVER_FEATURE_DMABUF = 1ull << 0, NEURON_DRIVER_FEATURE_ASYNC_DMA = 1ull << 1, @@ -20,6 +23,7 @@ enum neuron_driver_feature_flag { NEURON_DRIVER_FEATURE_ZEROCOPY = 1ull << 8, NEURON_DRIVER_FEATURE_PINNED_HOST_MEM = 1ull << 9, NEURON_DRIVER_FEATURE_ALLOC_WITH_PA = 1ull << 10, + NEURON_DRIVER_FEATURE_ASYNC_IO = 1ull << 11, }; // FIXME this should be more generic - like node type. diff --git a/v2/address_map.h b/v2/address_map.h index 5607b43..9c801f8 100644 --- a/v2/address_map.h +++ b/v2/address_map.h @@ -56,7 +56,7 @@ #define V2_TS_PER_DEVICE 6 // max channels supported by v2 device -#define V2_MAX_DRAM_CHANNELS 2 //2 HBM's. +#define V2_NUM_HBMS 2 //2 HBM's. #define V2_MAX_DDR_REGIONS V2_NC_PER_DEVICE #define V2_SEMAPHORE_COUNT 256 diff --git a/v2/neuron_dhal_v2.c b/v2/neuron_dhal_v2.c index f7c8de7..3266ca0 100644 --- a/v2/neuron_dhal_v2.c +++ b/v2/neuron_dhal_v2.c @@ -309,8 +309,7 @@ static void ts_nq_set_hwaddr_v2(struct neuron_device *nd, u8 ts_id, u8 index, u3 * @nq_type: type of the notification queue * @size: size of queue in bytes * @on_host_memory: if true, NQ is created in host memory - * @dram_channel: If NQ is created on device memory which DRAM channel to use. - * @dram_region: If NQ is created on device memory which DRAM region to use. + * @hbm_index: If NQ is created on device memory which HBM to use. * @force_alloc_mem: If true, force allocate new memory (and delete already allocated memory, if any) * @nq_mc[out]: memchunk used by the NQ will be written here * @mc_ptr[out]: Pointer to memchunk backing this NQ @@ -318,7 +317,7 @@ static void ts_nq_set_hwaddr_v2(struct neuron_device *nd, u8 ts_id, u8 index, u3 * Return: 0 on if initialization succeeds, a negative error code otherwise. */ static int ts_nq_init_v2(struct neuron_device *nd, u8 ts_id, u8 eng_index, u32 nq_type, u32 size, - u32 on_host_memory, u32 dram_channel, u32 dram_region, + u32 on_host_memory, u32 hbm_index, bool force_alloc_mem, struct mem_chunk **nq_mc, u64 *mmap_offset) { // Check that size is power of 2 @@ -331,7 +330,7 @@ static int ts_nq_init_v2(struct neuron_device *nd, u8 ts_id, u8 eng_index, u32 n return -EINVAL; u8 nq_id = ts_nq_get_nqid_v2(nd, eng_index, nq_type); - if (nq_id >= MAX_NQ_SUPPORTED) + if (nq_id >= V2_MAX_NQ_SUPPORTED) return -EINVAL; struct mem_chunk *mc = nd->ts_nq_mc[ts_id][nq_id]; @@ -339,7 +338,7 @@ static int ts_nq_init_v2(struct neuron_device *nd, u8 ts_id, u8 eng_index, u32 n struct mem_chunk *_mc = NULL; u32 nc_id = ts_id / (V2_TS_PER_DEVICE / V2_NC_PER_DEVICE); int ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, size, (on_host_memory) ? 0 : size, on_host_memory ? MEM_LOC_HOST : MEM_LOC_DEVICE, - dram_channel, dram_region, nc_id, on_host_memory ? NEURON_MEMALLOC_TYPE_NOTIFICATION_HOST : NEURON_MEMALLOC_TYPE_NOTIFICATION_DEVICE, &_mc); + hbm_index, nc_id, on_host_memory ? NEURON_MEMALLOC_TYPE_NOTIFICATION_HOST : NEURON_MEMALLOC_TYPE_NOTIFICATION_DEVICE, &_mc); if (ret) return ret; ts_nq_set_hwaddr_v2(nd, ts_id, eng_index, nq_type, size, _mc->pa); @@ -467,8 +466,7 @@ static void nnq_set_hwaddr_v2(struct neuron_device *nd, u8 nc_id, u8 index, u32 */ static void mpset_set_dram_and_mpset_info_v2(struct neuron_mempool_set *mpset, u64 *device_dram_addr, u64 *device_dram_size) { - mpset->num_channels = V2_MAX_DRAM_CHANNELS; - mpset->mp_device_num_regions = 1; + mpset->num_hbms = V2_NUM_HBMS; device_dram_addr[0] = V2_HBM_0_BASE; device_dram_addr[1] = V2_HBM_1_BASE; device_dram_size[0] = V2_HBM_0_SIZE; @@ -1400,7 +1398,7 @@ static void ndhal_ext_cleanup_v2(void) * static asserts to valid static const sizes work across versions * */ -static_assert( MAX_DRAM_CHANNELS >= V2_MAX_DRAM_CHANNELS, "Max dram channel count too small"); +static_assert( MAX_NUM_HBMS >= V2_NUM_HBMS, "Max dram channel count too small"); static_assert( MAX_TS_PER_DEVICE >= V2_TS_PER_DEVICE, "Max ts per device count too small"); static_assert( MAX_NC_PER_DEVICE >= V2_NC_PER_DEVICE, "Max nc per device count too small"); static_assert( MAX_NQ_TYPE >= V2_MAX_NQ_TYPE, "Max nq type count too small"); @@ -1435,7 +1433,8 @@ int ndhal_register_funcs_v2(void) { ndhal->ndhal_address_map.event_count = V2_EVENTS_COUNT; ndhal->ndhal_address_map.ts_per_device = V2_TS_PER_DEVICE; ndhal->ndhal_address_map.dma_eng_per_nc = V2_DMA_ENG_PER_NC; - ndhal->ndhal_address_map.dram_channels = V2_MAX_DRAM_CHANNELS; + ndhal->ndhal_address_map.num_hbms = V2_NUM_HBMS; + ndhal->ndhal_address_map.nq_per_nc = V2_MAX_NQ_SUPPORTED; ndhal->ndhal_reset.initiate_max_wait_time = V2_NR_RESET_INIT_MAX_TOTAL_WAIT_TIME_MS; ndhal->ndhal_reset.retry_count = NR_RESET_RETRY_COUNT; ndhal->ndhal_reset.nr_post_reset_config = nr_post_reset_config_v2; diff --git a/v3/address_map.h b/v3/address_map.h index d1c9ac1..ad073c1 100644 --- a/v3/address_map.h +++ b/v3/address_map.h @@ -71,7 +71,7 @@ #define V3_TS_PER_DEVICE (V3_NC_PER_DEVICE * V3_TS_PER_NC) // max channels supported by V3 device -#define V3_MAX_DRAM_CHANNELS 4 +#define V3_NUM_HBMS 4 #define V3_SEMAPHORE_COUNT 256 #define V3_EVENTS_COUNT 256 @@ -79,7 +79,7 @@ #define V3_ALLOWED_DESC_PER_PACKET 64 #define V3_MAX_NQ_QUEUES 16 -#define V3_MAX_NQ_TYPE 5 +#define V3_MAX_NQ_TYPE 6 #define V3_MAX_NQ_SUPPORTED (V3_MAX_NQ_TYPE * V3_MAX_NQ_QUEUES) #define V3_APB_IO_0_BASE 0x00008000000000ull diff --git a/v3/neuron_dhal_v3.c b/v3/neuron_dhal_v3.c index e70d99d..678e84c 100644 --- a/v3/neuron_dhal_v3.c +++ b/v3/neuron_dhal_v3.c @@ -233,32 +233,22 @@ static int ndhal_register_funcs_trn2(void) { return 0; } -/* Instance names - */ -#define NEURON_TRN2P_INSTANCE_NAME "trn2p.48xlarge" -#define NEURON_TRN2EU_INSTANCE_NAME "trn2eu.48xlarge" -#define NEURON_TRN2U_INSTANCE_NAME "trn2u.48xlarge" -#define NEURON_TRN2UAC_INSTANCE_NAME "trn2u-ac.24xlarge" -#define NEURON_TRN2PDS_INSTANCE_NAME "trn2es.48xlarge" +static const struct neuron_platform_lookup v3_platform_map[] = { + {"trn2p.48xlarge", NEURON_PLATFORM_TYPE_ULTRASERVER}, + {"trn2eu.48xlarge", NEURON_PLATFORM_TYPE_ULTRASERVER}, + {"trn2u.48xlarge", NEURON_PLATFORM_TYPE_ULTRASERVER}, + {"trn2u-ac.24xlarge", NEURON_PLATFORM_TYPE_ULTRASERVER}, + {"trn2es.48xlarge", NEURON_PLATFORM_TYPE_PDS}, + {"trn2en.24xlarge", NEURON_PLATFORM_TYPE_MAX}, + {NULL, NEURON_PLATFORM_TYPE_INVALID}, +}; static enum neuron_platform_type ndhal_platform_type_v3(void) { - enum neuron_platform_type platform_type = NEURON_PLATFORM_TYPE_INVALID; - char buf[128]; + enum neuron_platform_type platform_type; - if (narch_get_instance_type_name(buf, sizeof(buf))) goto done; - if ((strncmp(buf, NEURON_TRN2P_INSTANCE_NAME, sizeof(NEURON_TRN2P_INSTANCE_NAME)-1) == 0) || - (strncmp(buf, NEURON_TRN2EU_INSTANCE_NAME, sizeof(NEURON_TRN2EU_INSTANCE_NAME)-1) == 0) || - (strncmp(buf, NEURON_TRN2U_INSTANCE_NAME, sizeof(NEURON_TRN2U_INSTANCE_NAME)-1) == 0) || - (strncmp(buf, NEURON_TRN2UAC_INSTANCE_NAME, sizeof(NEURON_TRN2UAC_INSTANCE_NAME)-1) == 0)) { - platform_type = NEURON_PLATFORM_TYPE_ULTRASERVER; - } else if ((strncmp(buf, NEURON_TRN2PDS_INSTANCE_NAME, sizeof(NEURON_TRN2PDS_INSTANCE_NAME)-1) == 0)) { - platform_type = NEURON_PLATFORM_TYPE_PDS; - } else { - platform_type = NEURON_PLATFORM_TYPE_STD; - } + platform_type = narch_detect_platform_type(v3_platform_map); -done: if (force_userver) { platform_type = NEURON_PLATFORM_TYPE_ULTRASERVER; } @@ -506,8 +496,7 @@ static void ts_nq_set_hwaddr_v3(struct neuron_device *nd, u8 ts_id, u8 index, u3 * @nq_type: type of the notification queue * @size: size of queue in bytes * @on_host_memory: if true, NQ is created in host memory - * @dram_channel: If NQ is created on device memory which DRAM channel to use. - * @dram_region: If NQ is created on device memory which DRAM region to use. + * @hbm_index: If NQ is created on device memory which HBM to use. * @force_alloc_mem: If true, force allocate new memory (and delete already allocated memory, if any) * @nq_mc[out]: memchunk used by the NQ will be written here * @mc_ptr[out]: Pointer to memchunk backing this NQ @@ -515,7 +504,7 @@ static void ts_nq_set_hwaddr_v3(struct neuron_device *nd, u8 ts_id, u8 index, u3 * Return: 0 on if initialization succeeds, a negative error code otherwise. */ static int ts_nq_init_v3(struct neuron_device *nd, u8 ts_id, u8 eng_index, u32 nq_type, u32 size, - u32 on_host_memory, u32 dram_channel, u32 dram_region, + u32 on_host_memory, u32 hbm_index, bool force_alloc_mem, struct mem_chunk **nq_mc, u64 *mmap_offset) { // Check that size is power of 2 @@ -528,7 +517,7 @@ static int ts_nq_init_v3(struct neuron_device *nd, u8 ts_id, u8 eng_index, u32 n return -EINVAL; u8 nq_id = ts_nq_get_nqid_v3(nd, eng_index, nq_type); - if (nq_id >= MAX_NQ_SUPPORTED) + if (nq_id >= V3_MAX_NQ_SUPPORTED) return -EINVAL; struct mem_chunk *mc = nd->ts_nq_mc[ts_id][nq_id]; @@ -537,7 +526,7 @@ static int ts_nq_init_v3(struct neuron_device *nd, u8 ts_id, u8 eng_index, u32 n u32 nc_id = ts_id / V3_TS_PER_NC; int ret = mc_alloc_align(nd, MC_LIFESPAN_DEVICE, size, (on_host_memory) ? 0 : size, on_host_memory ? MEM_LOC_HOST : MEM_LOC_DEVICE, - dram_channel, dram_region, nc_id, on_host_memory ? NEURON_MEMALLOC_TYPE_NOTIFICATION_HOST : NEURON_MEMALLOC_TYPE_NOTIFICATION_DEVICE, &_mc); + hbm_index, nc_id, on_host_memory ? NEURON_MEMALLOC_TYPE_NOTIFICATION_HOST : NEURON_MEMALLOC_TYPE_NOTIFICATION_DEVICE, &_mc); if (ret) return ret; ts_nq_set_hwaddr_v3(nd, ts_id, eng_index, nq_type, size, _mc->pa); @@ -664,8 +653,7 @@ static void nnq_set_hwaddr_v3(struct neuron_device *nd, u8 nc_id, u8 index, u32 */ static void mpset_set_dram_and_mpset_info_v3(struct neuron_mempool_set *mpset, u64 *device_dram_addr, u64 *device_dram_size) { - mpset->num_channels = V3_MAX_DRAM_CHANNELS; - mpset->mp_device_num_regions = 1; + mpset->num_hbms = V3_NUM_HBMS; device_dram_addr[0] = V3_HBM_0_BASE; device_dram_addr[1] = V3_HBM_1_BASE; device_dram_addr[2] = V3_HBM_2_BASE; @@ -697,7 +685,7 @@ static void mpset_set_dram_and_mpset_info_v3(struct neuron_mempool_set *mpset, u device_dram_size[3] = V3_HBM_ACTIVE_SIZE; } int i; - for (i = 0; i < mpset->num_channels; i++) { + for (i = 0; i < mpset->num_hbms; i++) { ndhal->ndhal_mpset.device_dram_end_addr[i] = device_dram_addr[i] + device_dram_size[i]; } } @@ -1166,16 +1154,25 @@ static const u32 v3_pds_routing_id_to_user_id[] = { 12, 13, 14, 15 }; +static const u32 v3_max_routing_id_to_user_id[] = { + 0, 1, 2, 3 }; + #define V3_ROUTING_ID_TBL_SZ (sizeof(v3_torus_routing_id_to_user_id) / sizeof(v3_torus_routing_id_to_user_id[0])) +#define V3_MAX_ROUTING_ID_TBL_SZ (sizeof(v3_max_routing_id_to_user_id) / sizeof(v3_max_routing_id_to_user_id[0])) static u32 neuron_pci_routing_id_to_user_id(u32 routing_id) { if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { return v3_pds_routing_id_to_user_id[ routing_id % V3_ROUTING_ID_TBL_SZ]; } + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { + return v3_max_routing_id_to_user_id[ routing_id % V3_MAX_ROUTING_ID_TBL_SZ]; + } return v3_torus_routing_id_to_user_id[ routing_id % V3_ROUTING_ID_TBL_SZ]; } +static atomic_t v3_trn2max_device_count = ATOMIC_INIT(0); + /** * neuron_pci_get_device_id() - get device id and set nd->device_index * @@ -1190,6 +1187,14 @@ static int neuron_pci_get_device_id_v3(struct neuron_device *nd, struct pci_dev u32 routing_id = (u32)-1; u32 routing_id_max = MAX_NEURON_DEVICE_COUNT; + // FW does not populate routing_id miscram regs on Trn2MAX servers, + // so assign device IDs sequentially by probe order instead. + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { + routing_id = (u32)dhal_atomic_fetch_add(1, &v3_trn2max_device_count); + nd->device_index = neuron_pci_routing_id_to_user_id(routing_id); + goto done; + } + // Poll the device id until the device is ready for (i = 0; i < 20; i++) { ret = fw_io_device_id_read(nd->npdev.bar0, &routing_id); @@ -1228,6 +1233,7 @@ static int neuron_pci_get_device_id_v3(struct neuron_device *nd, struct pci_dev nd->device_index = neuron_pci_routing_id_to_user_id(routing_id); +done: pr_err("** BDF: %2.2x:%2.2x.%x => nd[%d] (routing id: %u)\n", dev->bus->number, PCI_SLOT(dev->devfn), PCI_FUNC(dev->devfn), nd->device_index, routing_id); // protection against duplicate IDs - doesn't provide 100% protection in multi-threaded device discovery @@ -1252,25 +1258,14 @@ neuron_pci_device_id_to_rid_map_v3(uint32_t * count, uint32_t * did_to_rid_map) { int i; - switch (ndhal->pci_device_id) { - case TRN2_DEVICE_ID0: - case TRN3_DEVICE_ID0: - case TRN3_DEVICE_ID1: - for (i = 0; i < total_neuron_devices; i++) { - u32 routing_id; - if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { - routing_id = i + ndhal->ndhal_arch.server_id * total_neuron_devices; - } else { - routing_id = i; - } - did_to_rid_map[neuron_pci_routing_id_to_user_id(routing_id)] = routing_id; - } - break; - - default: - for (i = 0; i < total_neuron_devices; i++) { - did_to_rid_map[i] = i; - } + for (i = 0; i < total_neuron_devices; i++) { + u32 routing_id; + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + routing_id = i + ndhal->ndhal_arch.server_id * total_neuron_devices; + } else { + routing_id = i; + } + did_to_rid_map[neuron_pci_routing_id_to_user_id(routing_id)] = routing_id; } *count = total_neuron_devices; @@ -1691,7 +1686,8 @@ static int npe_pod_info_v3(u8 *pod_type, u8 *pod_id, u8 *pod_sz, enum neuron_ult *pod_sz = 0; *mode = NEURON_ULTRASERVER_MODE_UNSET; *modes_supported = 0; - } else if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + } else if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { *pod_type = NEURON_POD_TYPE_SWITCH; npe_get_pod_sz(pod_sz); npe_get_pod_id(pod_id); @@ -1826,7 +1822,8 @@ static ssize_t npe_class_node_id_show_data_v3(char *buf, u32 sz) */ static ssize_t npe_class_node_cnt_show_data_v3(char *buf) { - if (ndhal->ndhal_arch.platform_type != NEURON_PLATFORM_TYPE_PDS) { + if (ndhal->ndhal_arch.platform_type != NEURON_PLATFORM_TYPE_PDS && + ndhal->ndhal_arch.platform_type != NEURON_PLATFORM_TYPE_MAX) { return dhal_sysfs_emit(buf, "-1\n"); } return npe_class_node_cnt_show_data(buf); @@ -1907,7 +1904,7 @@ static void ndhal_ext_cleanup_v3(void) * static asserts to valid static const sizes work across versions * */ -static_assert( MAX_DRAM_CHANNELS >= V3_MAX_DRAM_CHANNELS, "Max dram channel count too small"); +static_assert( MAX_NUM_HBMS >= V3_NUM_HBMS, "Max dram channel count too small"); static_assert( MAX_TS_PER_DEVICE >= V3_TS_PER_DEVICE, "Max ts per device count too small"); static_assert( MAX_NC_PER_DEVICE >= V3_NC_PER_DEVICE, "Max nc per device count too small"); static_assert( MAX_NQ_TYPE >= V3_MAX_NQ_TYPE, "Max nq type count too small"); @@ -1943,7 +1940,8 @@ int ndhal_register_funcs_v3(void) { ndhal->ndhal_address_map.event_count = V3_EVENTS_COUNT; ndhal->ndhal_address_map.ts_per_device = V3_TS_PER_DEVICE; ndhal->ndhal_address_map.dma_eng_per_nc = V3_DMA_ENG_PER_NC; - ndhal->ndhal_address_map.dram_channels = V3_MAX_DRAM_CHANNELS; + ndhal->ndhal_address_map.num_hbms = V3_NUM_HBMS; + ndhal->ndhal_address_map.nq_per_nc = V3_MAX_NQ_SUPPORTED; ndhal->ndhal_reset.initiate_max_wait_time = V3_NR_RESET_INIT_MAX_TOTAL_WAIT_TIME_MS; ndhal->ndhal_reset.retry_count = NR_RESET_RETRY_COUNT; ndhal->ndhal_reset.nr_post_reset_config = nr_post_reset_config_v3; diff --git a/v3/neuron_pelect.c b/v3/neuron_pelect.c index 96b2949..84c971f 100644 --- a/v3/neuron_pelect.c +++ b/v3/neuron_pelect.c @@ -292,7 +292,7 @@ typedef struct pod_neighbor_io { struct mem_chunk *data_mc; } pod_neighbor_io_t; -static void npe_pds_config_init(void); +void npe_pds_config_init(void); static bool npe_pod_ctl_is_set(int value) { @@ -341,19 +341,19 @@ static int npe_pod_neighbor_io_init(pod_neighbor_io_t* pnio, struct neuron_devic goto done; } - ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, pnio->ring_size * sizeof(union udma_desc), 0, MEM_LOC_HOST, 0, 0, 0, NEURON_MEMALLOC_TYPE_MISC_HOST, &pnio->tx_mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, pnio->ring_size * sizeof(union udma_desc), 0, MEM_LOC_HOST, 0, 0, NEURON_MEMALLOC_TYPE_MISC_HOST, &pnio->tx_mc); if (ret) { pr_err("ultraserver election io memory allocation failed"); goto done; } - ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, pnio->ring_size * sizeof(union udma_desc), 0, MEM_LOC_HOST, 0, 0, 0, NEURON_MEMALLOC_TYPE_MISC_HOST, &pnio->rx_mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, pnio->ring_size * sizeof(union udma_desc), 0, MEM_LOC_HOST, 0, 0, NEURON_MEMALLOC_TYPE_MISC_HOST, &pnio->rx_mc); if (ret) { pr_err("ultraserver election io memory allocation failed"); goto done; } - ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, pnio->data_size, 0, MEM_LOC_HOST, 0, 0, 0, NEURON_MEMALLOC_TYPE_MISC_HOST, &pnio->data_mc); + ret = mc_alloc_align(nd, MC_LIFESPAN_LOCAL, pnio->data_size, 0, MEM_LOC_HOST, 0, 0, NEURON_MEMALLOC_TYPE_MISC_HOST, &pnio->data_mc); if (ret) { pr_err("ultraserver election io memory allocation failed"); goto done; @@ -1130,7 +1130,8 @@ static void npe_initiate_election(u64 nbr_data_read_timeout) static bool npe_all_rst_complete(void) { int i; - for (i=0; i < 16; i++) { + int device_count = (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) ? 4 : 16; + for (i = 0; i < device_count; i++) { if (ndhal_pelect_data.pnd[i] == NULL) { return false; } @@ -1212,6 +1213,7 @@ int npe_election_exec_on_rst(struct neuron_device *nd, bool reset_successful) // if ((ndhal_pelect_data.pod_state_internal != NEURON_NPE_POD_ST_INIT) || (ndhal->ndhal_arch.platform_type != NEURON_PLATFORM_TYPE_PDS && + ndhal->ndhal_arch.platform_type != NEURON_PLATFORM_TYPE_MAX && npe_pod_ctl_is_set(NPE_POD_CTL_RST_SKIP_ELECTION))) { goto done; } @@ -1222,9 +1224,8 @@ int npe_election_exec_on_rst(struct neuron_device *nd, bool reset_successful) goto done; } - // initialize PDS configuration (topology/election) data - // - if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { npe_pds_config_init(); goto done; } @@ -1278,8 +1279,9 @@ static int npe_get_modal_node_id(enum neuron_ultraserver_mode mode) { int node_id = ndhal_pelect_data.node_id; - // PDS doesn't change node_id based on mode because nodes id are location based vs. election based - if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + // PDS/MAX doesn't change node_id based on mode because nodes id are location based vs. election based + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { return node_id; } @@ -1620,8 +1622,9 @@ int npe_pod_ctrl(struct neuron_device *nd, u32 ctrl, enum neuron_ultraserver_mod } else if (ctrl == NEURON_NPE_POD_CTRL_REQ_POD) { int mark_cnt = ncrwl_range_mark_cnt_get(); - // no election required on PDS, return success - if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + // no election required on PDS/MAX, return success + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { ret = 0; goto done; } @@ -1657,7 +1660,8 @@ int npe_pod_ctrl(struct neuron_device *nd, u32 ctrl, enum neuron_ultraserver_mod */ int npe_platform_ready(struct neuron_device *nd, enum neuron_platform_operation_type platform_operation) { - if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { switch (platform_operation) { case NEURON_PLATFORM_OP_TYPE_DEVOPEN: if (npe_pod_state_busy()) { @@ -1810,7 +1814,8 @@ ssize_t npe_class_node_id_show_data(char *buf, u32 sz) return dhal_sysfs_emit(buf, "busy\n"); } - if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { mode = npe_node_cnt_to_mode(ndhal_pelect_data.node_cnt); } else if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_ULTRASERVER) { mode = npe_node_cnt_to_mode(sz); @@ -1825,9 +1830,10 @@ ssize_t npe_class_node_id_show_data(char *buf, u32 sz) ssize_t npe_class_node_cnt_show_data(char *buf) { - int node_cnt = -1; // node_cnt is currently only returned for PDS + int node_cnt = -1; // node_cnt is currently only returned for PDS/MAX - if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { node_cnt = ndhal_pelect_data.node_cnt; } @@ -1849,7 +1855,8 @@ ssize_t npe_class_server_id_show_data(char *buf, u32 sz) return dhal_sysfs_emit(buf, "0000000000000000\n"); } - if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { mode = npe_node_cnt_to_mode(ndhal_pelect_data.node_cnt); } else if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_ULTRASERVER) { mode = npe_node_cnt_to_mode(sz); @@ -1911,7 +1918,7 @@ struct { * - node_cnt - count of nodes in the PDS server * */ -static void npe_pds_config_init(void) +void npe_pds_config_init(void) { static bool initialized = false; int ret = 0; @@ -1927,8 +1934,6 @@ static void npe_pds_config_init(void) return; } - // first check the mapping table to see if there's a match for preview server - // nd = ndhal_pelect_data.pnd[0]; if (nd == NULL) { pr_err("internal error. Neuron device pointer should not be null at start of election"); @@ -1962,8 +1967,9 @@ static void npe_pds_config_init(void) } if ((partition_sz == -1) || (instance_sz <= 0)) { - pr_warn("PDS partition/instance size data is invalid (%d/%d), defaulting to 4 node PDS configuration", partition_sz, instance_sz); - ndhal_pelect_data.node_cnt = 4; + int default_node_cnt = (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) ? 1 : 4; + pr_warn("partition/instance size data is invalid (%d/%d), defaulting to %d node configuration", partition_sz, instance_sz, default_node_cnt); + ndhal_pelect_data.node_cnt = default_node_cnt; } else { ndhal_pelect_data.node_cnt = partition_sz / instance_sz; } diff --git a/v3/neuron_pelect.h b/v3/neuron_pelect.h index 06f87ff..0f99af0 100644 --- a/v3/neuron_pelect.h +++ b/v3/neuron_pelect.h @@ -25,6 +25,11 @@ int npe_election_exec_on_rst(struct neuron_device *nd, bool reset_successful); */ int npe_init(void); +/** + * npe_pds_config_init() - initialize PDS/MAX platform topology from firmware + */ +void npe_pds_config_init(void); + /** * npe_cleanup() - cleanup and pod state left around (miscram) * diff --git a/v4/address_map.h b/v4/address_map.h index 401925e..225e914 100644 --- a/v4/address_map.h +++ b/v4/address_map.h @@ -71,7 +71,7 @@ #define V4_TS_PER_DEVICE (V4_NC_PER_DEVICE * V4_TS_PER_NC) // max channels supported by V4 device -#define V4_MAX_DRAM_CHANNELS 4 +#define V4_NUM_HBMS 4 #define V4_SEMAPHORE_COUNT 256 #define V4_EVENTS_COUNT 256 @@ -79,7 +79,7 @@ #define V4_ALLOWED_DESC_PER_PACKET 64 #define V4_MAX_NQ_QUEUES 16 -#define V4_MAX_NQ_TYPE 5 +#define V4_MAX_NQ_TYPE 6 #define V4_MAX_NQ_SUPPORTED (V4_MAX_NQ_TYPE * V4_MAX_NQ_QUEUES) #define V4_APB_IO_0_BASE 0x00008000000000ull diff --git a/v4/neuron_dhal_v4.c b/v4/neuron_dhal_v4.c index 9ad23df..6dadfde 100644 --- a/v4/neuron_dhal_v4.c +++ b/v4/neuron_dhal_v4.c @@ -147,35 +147,25 @@ static int ndhal_register_funcs_trn3(void) { return 0; } -/* Instance names - */ -#define NEURON_TRN3PDS_INSTANCE_NAME "trn3.48xlarge" -#define NEURON_TRN3PDS0_INSTANCE_NAME "trn3-dev0.48xlarge" -#define NEURON_TRN3PDS1_INSTANCE_NAME "trn3-dev1.48xlarge" -#define NEURON_TRN3P_INSTANCE_NAME "trn3p.48xlarge" +static const struct neuron_platform_lookup v4_platform_map[] = { + {"trn3e.24xlarge", NEURON_PLATFORM_TYPE_MAX}, + {"trn3e-dev0.24xlarge", NEURON_PLATFORM_TYPE_MAX}, + {"trn3.48xlarge", NEURON_PLATFORM_TYPE_PDS}, + {"trn3-dev0.48xlarge", NEURON_PLATFORM_TYPE_PDS}, + {"trn3-dev1.48xlarge", NEURON_PLATFORM_TYPE_PDS}, + {"trn3p.48xlarge", NEURON_PLATFORM_TYPE_ULTRASERVER}, + {NULL, NEURON_PLATFORM_TYPE_INVALID}, +}; static enum neuron_platform_type ndhal_platform_type_v4(void) { - enum neuron_platform_type platform_type = NEURON_PLATFORM_TYPE_INVALID; - char buf[128]; + enum neuron_platform_type platform_type; - if (narch_get_instance_type_name(buf, sizeof(buf))) goto done; - if ((strncmp(buf, NEURON_TRN3PDS_INSTANCE_NAME, sizeof(NEURON_TRN3PDS_INSTANCE_NAME)-1) == 0)) { - platform_type = NEURON_PLATFORM_TYPE_PDS; - } else if ((strncmp(buf, NEURON_TRN3PDS0_INSTANCE_NAME, sizeof(NEURON_TRN3PDS0_INSTANCE_NAME)-1) == 0)) { - platform_type = NEURON_PLATFORM_TYPE_PDS; - } else if ((strncmp(buf, NEURON_TRN3PDS1_INSTANCE_NAME, sizeof(NEURON_TRN3PDS1_INSTANCE_NAME)-1) == 0)) { - platform_type = NEURON_PLATFORM_TYPE_PDS; - } else if ((strncmp(buf, NEURON_TRN3P_INSTANCE_NAME, sizeof(NEURON_TRN3P_INSTANCE_NAME)-1) == 0)) { - platform_type = NEURON_PLATFORM_TYPE_ULTRASERVER; - } else { - platform_type = NEURON_PLATFORM_TYPE_STD; - } + platform_type = narch_detect_platform_type(v4_platform_map); if (narch_is_qemu() || narch_is_emu()) platform_type = NEURON_PLATFORM_TYPE_STD; -done: return platform_type; } @@ -232,8 +222,7 @@ static void ndmar_ctx_queue_from_bit_v4(int bit, uint32_t *h2d_eng_id, uint32_t */ static void mpset_set_dram_and_mpset_info_v4(struct neuron_mempool_set *mpset, u64 *device_dram_addr, u64 *device_dram_size) { - mpset->num_channels = V4_MAX_DRAM_CHANNELS; - mpset->mp_device_num_regions = 1; + mpset->num_hbms = V4_NUM_HBMS; device_dram_addr[0] = V4_HBM_0_BASE; device_dram_addr[1] = V4_HBM_1_BASE; device_dram_addr[2] = V4_HBM_2_BASE; @@ -265,7 +254,7 @@ static void mpset_set_dram_and_mpset_info_v4(struct neuron_mempool_set *mpset, u device_dram_size[3] = V4_HBM_ACTIVE_SIZE; } int i; - for (i = 0; i < mpset->num_channels; i++) { + for (i = 0; i < mpset->num_hbms; i++) { ndhal->ndhal_mpset.device_dram_end_addr[i] = device_dram_addr[i] + device_dram_size[i]; } } @@ -323,13 +312,20 @@ static const u32 v4_pds_routing_id_to_user_id[] = { 12, 13, 14, 15 }; +static const u32 v4_max_routing_id_to_user_id[] = { + 0, 1, 2, 3 }; + #define V4_ROUTING_ID_TBL_SZ (sizeof(v4_torus_routing_id_to_user_id) / sizeof(v4_torus_routing_id_to_user_id[0])) +#define V4_MAX_ROUTING_ID_TBL_SZ (sizeof(v4_max_routing_id_to_user_id) / sizeof(v4_max_routing_id_to_user_id[0])) static u32 neuron_pci_routing_id_to_user_id(u32 routing_id) { if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { return v4_pds_routing_id_to_user_id[ routing_id % V4_ROUTING_ID_TBL_SZ]; } + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { + return v4_max_routing_id_to_user_id[ routing_id % V4_MAX_ROUTING_ID_TBL_SZ]; + } return v4_torus_routing_id_to_user_id[routing_id % V4_ROUTING_ID_TBL_SZ]; } @@ -367,7 +363,8 @@ static int neuron_pci_get_device_id_v4(struct neuron_device *nd, struct pci_dev return -ENODEV; } - if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { int server_id; ret = fw_io_server_info_read(nd->npdev.bar0, &server_id, NULL); @@ -453,6 +450,25 @@ static void perf_update_hbm_7200_supported_v4(struct neuron_device *nd) { return; } +static int neuron_pci_device_id_to_rid_map_v4(uint32_t *count, uint32_t *did_to_rid_map) +{ + int i; + + for (i = 0; i < total_neuron_devices; i++) { + u32 routing_id; + if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS || + ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { + routing_id = i + ndhal->ndhal_arch.server_id * total_neuron_devices; + } else { + routing_id = i; + } + did_to_rid_map[neuron_pci_routing_id_to_user_id(routing_id)] = routing_id; + } + + *count = total_neuron_devices; + return 0; +} + /** * ndhal_register_funcs_v4() - initialize the dhal for v4 chips * @@ -470,6 +486,7 @@ int ndhal_register_funcs_v4(void) { ndhal->ndhal_arch.platform_type = ndhal_platform_type_v4(); ndhal->ndhal_fw_io.new_readless_read_min_api_version = 6; ndhal->ndhal_pci.neuron_pci_get_device_id = neuron_pci_get_device_id_v4; + ndhal->ndhal_pci.neuron_pci_device_id_to_rid_map = neuron_pci_device_id_to_rid_map_v4; ndhal->ndhal_npe.npe_neighbor_eng_ids = npe_neighbor_eng_ids_v4; ndhal->ndhal_mpset.mpset_set_dram_and_mpset_info = mpset_set_dram_and_mpset_info_v4; ndhal->ndhal_mmap.dm_mmap_special = dm_mmap_special_v4; @@ -497,6 +514,8 @@ int ndhal_register_funcs_v4(void) { } else if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_PDS) { //TODO PDS ndhal->ndhal_cdev.ncdev_logical_to_physical_nc_map = ncdev_logical_to_physical_nc_map_v4; + } else if (ndhal->ndhal_arch.platform_type == NEURON_PLATFORM_TYPE_MAX) { + ndhal->ndhal_cdev.ncdev_logical_to_physical_nc_map = ncdev_logical_to_physical_nc_map_v4; } switch (ndhal->pci_device_id) {