Rename gds backend to cufile and bump to v0.3.1

README.md

@@ -49,8 +49,8 @@ filperf <device-uri>[,<device-uri>,...] [options]
 | Option | Default | Description |
 |---|---|---|
 | `--data-dir <name>` | _(none)_ | Root directory name containing class subdirectories |
-| `--backend <name>` | `aisio-cpu` | I/O backend: `aisio-cpu`, `aisio-gpu`, `aisio-p2p`, `posix`, `gds` |
-| `--mnt <path>` | `/mnt` | Mountpoint of the drive (for `posix` and `gds` backends) |
+| `--backend <name>` | `aisio-cpu` | I/O backend: `aisio-cpu`, `aisio-gpu`, `aisio-p2p`, `posix`, `cufile` |
+| `--mnt <path>` | `/mnt` | Mountpoint of the drive (for `posix` and `cufile` backends) |
 | `--batch-size <n>` | `1` | Number of files per batch |
 | `--batches <n>` | `1` | Number of batches to read |
 | `--iosize <n>` | `4096` | Number of bytes per I/O (`aisio-cpu` and `aisio-gpu` only) |
@@ -59,7 +59,7 @@ filperf <device-uri>[,<device-uri>,...] [options]
 | `--max-file-size <n>` | _(required)_ | Max file size in bytes; required for the `aisio-cpu`/`aisio-gpu`/`aisio-p2p` backends to size the upcie heap |
 | `--warmup <n>` | `0` | Un-timed batches to run before starting the measurement window |
 | `--buffered` | off | Disable `O_DIRECT` when using `posix` backend |
-| `--async` | off | Use async API when using `gds` backend |
+| `--async` | off | Use async API when using `cufile` backend |
 | `--summary` | off | Print I/O and dataset statistics after completion |
 | `--help` | | Print usage |
 
@@ -88,8 +88,10 @@ Dataset stats:
 	Average size of files in the dataset (KiB): 109.862692
 ```
 
-`posix` and `gds` both require a block device path and the device mounted at `--mnt`.
-`gds` additionally requires CUDA with cuFile for direct NVMe-to-GPU transfers.
+`posix` and `cufile` both require a block device path and the device mounted at `--mnt`.
+`cufile` additionally requires CUDA with the cuFile library. Direct NVMe-to-GPU
+(GPUDirect Storage) transfers need the kernel-side GDS driver (`nvidia-fs`); without
+it, cuFile falls back to a compatibility path that stages through host memory.
 For these file-level backends, IOPS equals File/s since each file is a single I/O.
 
 ```
@@ -114,7 +116,7 @@ Dataset stats:
 ```
 
 ```
-$ filperf /dev/nvme0n1 --batches 10 --batch-size 1024 --backend gds --mnt /mnt/datasets --data-dir imagenetish --summary
+$ filperf /dev/nvme0n1 --batches 10 --batch-size 1024 --backend cufile --mnt /mnt/datasets --data-dir imagenetish --summary
 Time, Batches, IOPS, MiB/s
 1.255441, 4, 3262.598873, 348.563276
 2.413140, 8, 3538.085634, 381.345474

include/fil_io.h

@@ -28,7 +28,7 @@ struct fil_file_io {
 	void *buffer;
 };
 
-struct fil_gds_io {
+struct fil_cufile_io {
 	CUfileDescr_t *descr;
 	CUfileHandle_t *handle;
 	size_t *expected;
@@ -59,6 +59,6 @@ int
 fil_file_submit(struct fil_iter *iter);
 
 int
-fil_gds_async_submit(struct fil_iter *iter);
+fil_cufile_async_submit(struct fil_iter *iter);
 
 #endif

include/fil_iter.h

@@ -34,7 +34,7 @@ struct fil_iter {
 	struct fil_stats *stats;
 	struct fil_opts *opts;
 	struct fil_output *output;
-	struct fil_gds_io *gds_io;
+	struct fil_cufile_io *cufile_io;
 	int (*io_fn)(struct fil_iter *iter);
 	struct fil_time *time;
 	uint64_t buffer_size;

include/libfil.h

@@ -41,7 +41,7 @@ struct fil_opts {
 			      ///< aisio-p2p backends to size the upcie heap correctly.
 	uint32_t batch_size; ///< The number of files per batch
 	bool buffered;       ///< Whether to use O_DIRECT with POSIX
-	bool async;          ///< Whether to use async API with GDS
+	bool async;          ///< Whether to use async API with cuFile
 };
 
 /**

meson.build

@@ -1,5 +1,5 @@
 project('fil', 'c', 'cuda',
-  version: '0.3.0',
+  version: '0.3.1',
   meson_version: '>=0.55.0',
   default_options: [
     'c_std=gnu11',

python/pyproject.toml

@@ -4,4 +4,4 @@ build-backend = "setuptools.build_meta"
 
 [project]
 name = "fil"
-version = "0.3.0"
+version = "0.3.1"

src/cli.c

@@ -26,9 +26,9 @@ print_help(const char *name)
 		"\t \t \t | \t The data-dir should be a name of a directory, not a path\n");
 	fprintf(stderr, "\t \t \t | \t The name of the data-dir should be unique\n");
 	fprintf(stderr, "\t --backend \t | \t The backend to use for reading files (aisio-cpu "
-			"[default], aisio-gpu, aisio-p2p, posix, gds)\n");
+			"[default], aisio-gpu, aisio-p2p, posix, cufile)\n");
 	fprintf(stderr, "\t --mnt \t \t | \t The mountpoint of the drive (default = /mnt). Only "
-			"relevant for backends: 'posix' and 'gds'\n");
+			"relevant for backends: 'posix' and 'cufile'\n");
 	fprintf(stderr, "\t --iosize \t | \t The number of bytes per I/O (default = 4096). Only "
 			"relevant for backends: 'aisio-cpu' and 'aisio-gpu'\n");
 	fprintf(stderr, "\t --gpu-nqueues \t | \t The number of GPU queues to create (default = "
@@ -43,7 +43,7 @@ print_help(const char *name)
 	fprintf(stderr, "\t --warmup \t | \t The number of un-timed warmup batches to read before "
 			"the timed run (default = 0)\n");
 	fprintf(stderr, "\t --buffered \t | \t Don't open with O_DIRECT when using POSIX\n");
-	fprintf(stderr, "\t --async \t | \t Use the async API when using GDS\n");
+	fprintf(stderr, "\t --async \t | \t Use the async API when using cuFile\n");
 	fprintf(stderr, "\t --summary \t | \t Print IO and dataset stats\n");
 	fprintf(stderr, "\t --help \t | \t Print this message\n");
 }

src/io.c

@@ -370,9 +370,9 @@ fil_file_submit(struct fil_iter *iter)
 	char *prefix, *path;
 	int fd, flags;
 	ssize_t err, bytes_read;
-	bool is_gds = strcmp(iter->opts->backend, "gds") == 0;
+	bool is_cufile = strcmp(iter->opts->backend, "cufile") == 0;
 	flags = O_RDONLY;
-	if (is_gds || !iter->opts->buffered) {
+	if (is_cufile || !iter->opts->buffered) {
 		flags |= O_DIRECT;
 	}
 
@@ -397,7 +397,7 @@ fil_file_submit(struct fil_iter *iter)
 		strcat(path, file->name);
 
 		nbytes = file->size;
-		if (!is_gds && !iter->opts->buffered) {
+		if (!is_cufile && !iter->opts->buffered) {
 			// POSIX O_DIRECT requires aligned nbytes
 			nbytes = (1 + ((file->size - 1) / xal_blksize)) * xal_blksize;
 		}
@@ -409,7 +409,7 @@ fil_file_submit(struct fil_iter *iter)
 			return err;
 		}
 
-		if (is_gds) {
+		if (is_cufile) {
 			memset(&descr, 0, sizeof(CUfileDescr_t));
 			descr.handle.fd = fd;
 			descr.type = CU_FILE_HANDLE_TYPE_OPAQUE_FD;
@@ -426,7 +426,7 @@ fil_file_submit(struct fil_iter *iter)
 
 		clock_gettime(CLOCK_MONOTONIC_RAW, &start);
 		bytes_read = 0;
-		if (is_gds) {
+		if (is_cufile) {
 
 			bytes_read = cuFileRead(fh, buffer, nbytes, 0, 0);
 			if (bytes_read < 0) {
@@ -476,19 +476,19 @@ fil_file_submit(struct fil_iter *iter)
 }
 
 int
-fil_gds_async_submit(struct fil_iter *iter)
+fil_cufile_async_submit(struct fil_iter *iter)
 {
 	struct xal_inode *dir, *file;
 	struct timespec start, end;
-	struct fil_gds_io *gds_io;
+	struct fil_cufile_io *cufile_io;
 	CUfileError_t status;
 	uint32_t buf_id, dev_id;
 	void *buffer;
 	char *prefix, *path;
 	int fd, flags, err = 0;
 	off_t offset = 0;
 
-	gds_io = iter->gds_io;
+	cufile_io = iter->cufile_io;
 	clock_gettime(CLOCK_MONOTONIC_RAW, &start);
 	for (uint32_t i = 0; i < iter->opts->batch_size; i++) {
 		dev_id = i % iter->n_devs;
@@ -516,20 +516,21 @@ fil_gds_async_submit(struct fil_iter *iter)
 			return err;
 		}
 
-		memset(&gds_io->descr[i], 0, sizeof(CUfileDescr_t));
-		gds_io->descr[i].handle.fd = fd;
-		gds_io->descr[i].type = CU_FILE_HANDLE_TYPE_OPAQUE_FD;
-		status = cuFileHandleRegister(&gds_io->handle[i], &gds_io->descr[i]);
+		memset(&cufile_io->descr[i], 0, sizeof(CUfileDescr_t));
+		cufile_io->descr[i].handle.fd = fd;
+		cufile_io->descr[i].type = CU_FILE_HANDLE_TYPE_OPAQUE_FD;
+		status = cuFileHandleRegister(&cufile_io->handle[i], &cufile_io->descr[i]);
 		if (status.err != CU_FILE_SUCCESS) {
 			fprintf(stderr, "Could not register file, err: %d\n", status.err);
 			close(fd);
 			return status.err;
 		}
-		gds_io->expected[i] = file->size;
-		gds_io->actual[i] = 0;
+		cufile_io->expected[i] = file->size;
+		cufile_io->actual[i] = 0;
 
-		status = cuFileReadAsync(gds_io->handle[i], buffer, &gds_io->expected[i], &offset,
-					 &offset, &gds_io->actual[i], gds_io->streams[i]);
+		status = cuFileReadAsync(cufile_io->handle[i], buffer, &cufile_io->expected[i],
+					 &offset, &offset, &cufile_io->actual[i],
+					 cufile_io->streams[i]);
 		if (status.err != CU_FILE_SUCCESS) {
 			fprintf(stderr, "cuFileReadAsync failed, err: %d\n", status.err);
 			err = status.err;
@@ -542,19 +543,19 @@ fil_gds_async_submit(struct fil_iter *iter)
 
 	clock_gettime(CLOCK_MONOTONIC_RAW, &start);
 	for (uint32_t i = 0; i < iter->opts->batch_size; i++) {
-		err = cudaStreamSynchronize(gds_io->streams[i]);
+		err = cudaStreamSynchronize(cufile_io->streams[i]);
 		if (err) {
 			fprintf(stderr, "Could not synchronize CUDA Stream, err: %d\n", err);
 			goto teardown;
 		}
-		if (gds_io->actual[i] < 0) {
-			err = gds_io->actual[i];
+		if (cufile_io->actual[i] < 0) {
+			err = cufile_io->actual[i];
 			fprintf(stderr, "Reading failed, err: %d\n", err);
 			goto teardown;
 		}
-		if ((size_t)gds_io->actual[i] != gds_io->expected[i]) {
+		if ((size_t)cufile_io->actual[i] != cufile_io->expected[i]) {
 			fprintf(stderr, "Could not read entire file, expected: %lu, actual: %lu\n",
-				gds_io->expected[i], gds_io->actual[i]);
+				cufile_io->expected[i], cufile_io->actual[i]);
 			err = EIO;
 			goto teardown;
 		}
@@ -564,8 +565,8 @@ fil_gds_async_submit(struct fil_iter *iter)
 
 teardown:
 	for (uint32_t i = 0; i < iter->opts->batch_size; i++) {
-		fd = gds_io->descr[i].handle.fd;
-		cuFileHandleDeregister(gds_io->handle[i]);
+		fd = cufile_io->descr[i].handle.fd;
+		cuFileHandleDeregister(cufile_io->handle[i]);
 		close(fd);
 	}
 	return err;

src/iter.c

@@ -91,7 +91,7 @@ _xnvme_setup(struct fil_iter *iter, struct fil_dev *device, const char *uri)
 	} else if (strcmp(backend, "posix") == 0) {
 		opts.be = "linux";
 		iter->type = FIL_FILE;
-	} else if (strcmp(backend, "gds") == 0) {
+	} else if (strcmp(backend, "cufile") == 0) {
 		opts.be = "linux";
 		iter->type = FIL_FILE;
 		status = cuFileDriverOpen();
@@ -539,50 +539,50 @@ _alloc(struct fil_iter *iter, uint32_t n_buffers)
 	}
 
 	if (iter->opts->async) {
-		iter->gds_io = malloc(sizeof(struct fil_gds_io));
-		if (!iter->gds_io) {
+		iter->cufile_io = malloc(sizeof(struct fil_cufile_io));
+		if (!iter->cufile_io) {
 			err = errno;
-			fprintf(stderr, "Could not allocate GDS IO struct: %d\n", err);
+			fprintf(stderr, "Could not allocate cuFile IO struct: %d\n", err);
 			return err;
 		}
 
-		iter->gds_io->descr = malloc(sizeof(CUfileDescr_t) * iter->opts->batch_size);
-		if (!iter->gds_io->descr) {
+		iter->cufile_io->descr = malloc(sizeof(CUfileDescr_t) * iter->opts->batch_size);
+		if (!iter->cufile_io->descr) {
 			err = errno;
 			fprintf(stderr, "Could not allocate cuFile descriptors: %d\n", err);
 			return err;
 		}
 
-		iter->gds_io->handle = malloc(sizeof(CUfileHandle_t) * iter->opts->batch_size);
-		if (!iter->gds_io->handle) {
+		iter->cufile_io->handle = malloc(sizeof(CUfileHandle_t) * iter->opts->batch_size);
+		if (!iter->cufile_io->handle) {
 			err = errno;
 			fprintf(stderr, "Could not allocate cuFile handles: %d\n", err);
 			return err;
 		}
 
-		iter->gds_io->expected = malloc(sizeof(size_t) * iter->opts->batch_size);
-		if (!iter->gds_io->expected) {
+		iter->cufile_io->expected = malloc(sizeof(size_t) * iter->opts->batch_size);
+		if (!iter->cufile_io->expected) {
 			err = errno;
 			fprintf(stderr, "Could not allocate array of expected values: %d\n", err);
 			return err;
 		}
 
-		iter->gds_io->actual = malloc(sizeof(ssize_t) * iter->opts->batch_size);
-		if (!iter->gds_io->actual) {
+		iter->cufile_io->actual = malloc(sizeof(ssize_t) * iter->opts->batch_size);
+		if (!iter->cufile_io->actual) {
 			err = errno;
 			fprintf(stderr, "Could not allocate array of actual values: %d\n", err);
 			return err;
 		}
 
-		iter->gds_io->streams = malloc(sizeof(cudaStream_t) * iter->opts->batch_size);
-		if (!iter->gds_io->streams) {
+		iter->cufile_io->streams = malloc(sizeof(cudaStream_t) * iter->opts->batch_size);
+		if (!iter->cufile_io->streams) {
 			err = errno;
 			fprintf(stderr, "Could not allocate array of CUDA Streams: %d\n", err);
 			return err;
 		}
 
 		for (uint32_t i = 0; i < iter->opts->batch_size; i++) {
-			err = cudaStreamCreateWithFlags(&iter->gds_io->streams[i],
+			err = cudaStreamCreateWithFlags(&iter->cufile_io->streams[i],
 							cudaStreamNonBlocking);
 			if (err) {
 				fprintf(stderr, "Could not setup CUDA Stream, err: %d\n", err);
@@ -656,15 +656,15 @@ fil_term(struct fil_iter *iter)
 		free(device->buffers);
 		free(device);
 	}
-	if (iter->gds_io) {
-		free(iter->gds_io->descr);
-		free(iter->gds_io->handle);
-		free(iter->gds_io->expected);
-		free(iter->gds_io->actual);
+	if (iter->cufile_io) {
+		free(iter->cufile_io->descr);
+		free(iter->cufile_io->handle);
+		free(iter->cufile_io->expected);
+		free(iter->cufile_io->actual);
 		for (uint32_t i = 0; i < iter->opts->batch_size; i++) {
-			cudaStreamDestroy(iter->gds_io->streams[i]);
+			cudaStreamDestroy(iter->cufile_io->streams[i]);
 		}
-		free(iter->gds_io->streams);
+		free(iter->cufile_io->streams);
 	}
 	if (iter->data) {
 		free(iter->data->entries);
@@ -713,8 +713,8 @@ fil_init(struct fil_iter **iter, char **dev_uris, uint32_t n_devs, struct fil_op
 		return EINVAL;
 	}
 
-	if (opts->async && strcmp(opts->backend, "gds") != 0) {
-		fprintf(stderr, "opts->async == true is only compatible with GDS backend");
+	if (opts->async && strcmp(opts->backend, "cufile") != 0) {
+		fprintf(stderr, "opts->async == true is only compatible with cuFile backend");
 		return EINVAL;
 	}
 
@@ -808,7 +808,7 @@ fil_init(struct fil_iter **iter, char **dev_uris, uint32_t n_devs, struct fil_op
 			break;
 		case FIL_FILE:
 			if (_iter->opts->async) {
-				_iter->io_fn = fil_gds_async_submit;
+				_iter->io_fn = fil_cufile_async_submit;
 			} else {
 				_iter->io_fn = fil_file_submit;
 			}