203 lines
7.4 KiB
C++
203 lines
7.4 KiB
C++
static void print_stress_report(const struct stress_report *report, int device_index, int seconds) {
|
|
printf("device=%s\n", report->device);
|
|
printf("device_index=%d\n", device_index);
|
|
printf("compute_capability=%d.%d\n", report->cc_major, report->cc_minor);
|
|
printf("backend=%s\n", report->backend);
|
|
printf("duration_s=%d\n", seconds);
|
|
printf("buffer_mb=%d\n", report->buffer_mb);
|
|
printf("streams=%d\n", report->stream_count);
|
|
printf("iterations=%lu\n", report->iterations);
|
|
printf("checksum=%llu\n", (unsigned long long)report->checksum);
|
|
if (report->details[0] != '\0') {
|
|
printf("%s", report->details);
|
|
}
|
|
printf("status=OK\n");
|
|
}
|
|
|
|
int main(int argc, char **argv) {
|
|
int seconds = 5;
|
|
int size_mb = 64;
|
|
int device_index = 0;
|
|
const char *precision_filter = NULL; /* NULL = all; else block_label to match */
|
|
#if HAVE_CUBLASLT_HEADERS
|
|
const char *precision_plan = NULL;
|
|
const char *precision_plan_seconds = NULL;
|
|
#endif
|
|
for (int i = 1; i < argc; i++) {
|
|
if ((strcmp(argv[i], "--seconds") == 0 || strcmp(argv[i], "-t") == 0) && i + 1 < argc) {
|
|
seconds = atoi(argv[++i]);
|
|
} else if ((strcmp(argv[i], "--size-mb") == 0 || strcmp(argv[i], "-m") == 0) && i + 1 < argc) {
|
|
size_mb = atoi(argv[++i]);
|
|
} else if ((strcmp(argv[i], "--device") == 0 || strcmp(argv[i], "-d") == 0) && i + 1 < argc) {
|
|
device_index = atoi(argv[++i]);
|
|
} else if (strcmp(argv[i], "--precision") == 0 && i + 1 < argc) {
|
|
precision_filter = argv[++i];
|
|
} else if (strcmp(argv[i], "--precision-plan") == 0 && i + 1 < argc) {
|
|
#if HAVE_CUBLASLT_HEADERS
|
|
precision_plan = argv[++i];
|
|
#else
|
|
fprintf(stderr, "--precision-plan requires a build with cuBLASLt headers\n");
|
|
return 2;
|
|
#endif
|
|
} else if (strcmp(argv[i], "--precision-plan-seconds") == 0 && i + 1 < argc) {
|
|
#if HAVE_CUBLASLT_HEADERS
|
|
precision_plan_seconds = argv[++i];
|
|
#else
|
|
fprintf(stderr, "--precision-plan-seconds requires a build with cuBLASLt headers\n");
|
|
return 2;
|
|
#endif
|
|
} else {
|
|
fprintf(stderr,
|
|
"usage: %s [--seconds N] [--size-mb N] [--device N] [--precision int8|fp8|fp16|fp32|fp64|fp4] [--precision-plan p1,p2,...,mixed] [--precision-plan-seconds s1,s2,...]\n",
|
|
argv[0]);
|
|
return 2;
|
|
}
|
|
}
|
|
if (seconds <= 0) {
|
|
seconds = 5;
|
|
}
|
|
if (size_mb <= 0) {
|
|
size_mb = 64;
|
|
}
|
|
if (device_index < 0) {
|
|
device_index = 0;
|
|
}
|
|
|
|
struct cuda_api cuda;
|
|
if (!load_cuda(&cuda)) {
|
|
fprintf(stderr, "failed to load libcuda.so.1 or required Driver API symbols\n");
|
|
return 1;
|
|
}
|
|
load_symbol(cuda.lib, "cuGetErrorName", (void **)&cuda.cuGetErrorName);
|
|
load_symbol(cuda.lib, "cuGetErrorString", (void **)&cuda.cuGetErrorString);
|
|
|
|
if (!check_rc(&cuda, "cuInit", cuda.cuInit(0))) {
|
|
return 1;
|
|
}
|
|
|
|
int count = 0;
|
|
if (!check_rc(&cuda, "cuDeviceGetCount", cuda.cuDeviceGetCount(&count))) {
|
|
return 1;
|
|
}
|
|
if (count <= 0) {
|
|
fprintf(stderr, "no CUDA devices found\n");
|
|
return 1;
|
|
}
|
|
|
|
if (device_index >= count) {
|
|
fprintf(stderr, "device index %d out of range (found %d CUDA device(s))\n", device_index, count);
|
|
return 1;
|
|
}
|
|
|
|
CUdevice dev = 0;
|
|
if (!check_rc(&cuda, "cuDeviceGet", cuda.cuDeviceGet(&dev, device_index))) {
|
|
return 1;
|
|
}
|
|
|
|
char name[128] = {0};
|
|
if (!check_rc(&cuda, "cuDeviceGetName", cuda.cuDeviceGetName(name, (int)sizeof(name), dev))) {
|
|
return 1;
|
|
}
|
|
|
|
int cc_major = 0;
|
|
int cc_minor = 0;
|
|
if (!query_compute_capability(&cuda, dev, &cc_major, &cc_minor)) {
|
|
return 1;
|
|
}
|
|
|
|
struct stress_report report;
|
|
int ok = 0;
|
|
|
|
#if HAVE_CUBLASLT_HEADERS
|
|
if (precision_plan != NULL && precision_plan[0] != '\0') {
|
|
char *plan_copy = strdup(precision_plan);
|
|
char *plan_seconds_copy = NULL;
|
|
int phase_seconds[32] = {0};
|
|
int phase_seconds_count = 0;
|
|
int phase_ok = 0;
|
|
if (plan_copy == NULL) {
|
|
fprintf(stderr, "failed to allocate precision plan buffer\n");
|
|
return 1;
|
|
}
|
|
if (precision_plan_seconds != NULL && precision_plan_seconds[0] != '\0') {
|
|
plan_seconds_copy = strdup(precision_plan_seconds);
|
|
if (plan_seconds_copy == NULL) {
|
|
free(plan_copy);
|
|
fprintf(stderr, "failed to allocate precision plan seconds buffer\n");
|
|
return 1;
|
|
}
|
|
for (char *sec_token = strtok(plan_seconds_copy, ",");
|
|
sec_token != NULL && phase_seconds_count < (int)(sizeof(phase_seconds) / sizeof(phase_seconds[0]));
|
|
sec_token = strtok(NULL, ",")) {
|
|
while (*sec_token == ' ' || *sec_token == '\t') {
|
|
sec_token++;
|
|
}
|
|
if (*sec_token == '\0') {
|
|
continue;
|
|
}
|
|
phase_seconds[phase_seconds_count++] = atoi(sec_token);
|
|
}
|
|
}
|
|
int phase_idx = 0;
|
|
for (char *token = strtok(plan_copy, ","); token != NULL; token = strtok(NULL, ","), phase_idx++) {
|
|
while (*token == ' ' || *token == '\t') {
|
|
token++;
|
|
}
|
|
if (*token == '\0') {
|
|
continue;
|
|
}
|
|
const char *phase_name = token;
|
|
const char *phase_filter = token;
|
|
if (strcmp(token, "mixed") == 0 || strcmp(token, "all") == 0) {
|
|
phase_filter = NULL;
|
|
}
|
|
int phase_duration = seconds;
|
|
if (phase_idx < phase_seconds_count && phase_seconds[phase_idx] > 0) {
|
|
phase_duration = phase_seconds[phase_idx];
|
|
}
|
|
printf("phase_begin=%s\n", phase_name);
|
|
fflush(stdout);
|
|
memset(&report, 0, sizeof(report));
|
|
ok = run_cublaslt_stress(&cuda, dev, name, cc_major, cc_minor, phase_duration, size_mb, phase_filter, &report);
|
|
if (ok) {
|
|
print_stress_report(&report, device_index, phase_duration);
|
|
phase_ok = 1;
|
|
} else {
|
|
printf("phase_error=%s\n", phase_name);
|
|
if (report.details[0] != '\0') {
|
|
printf("%s", report.details);
|
|
if (report.details[strlen(report.details) - 1] != '\n') {
|
|
printf("\n");
|
|
}
|
|
}
|
|
printf("status=FAILED\n");
|
|
}
|
|
printf("phase_end=%s\n", phase_name);
|
|
fflush(stdout);
|
|
}
|
|
free(plan_seconds_copy);
|
|
free(plan_copy);
|
|
return phase_ok ? 0 : 1;
|
|
}
|
|
ok = run_cublaslt_stress(&cuda, dev, name, cc_major, cc_minor, seconds, size_mb, precision_filter, &report);
|
|
#endif
|
|
if (!ok) {
|
|
if (precision_filter != NULL) {
|
|
fprintf(stderr,
|
|
"requested precision path unavailable: precision=%s device=%s cc=%d.%d\n",
|
|
precision_filter,
|
|
name,
|
|
cc_major,
|
|
cc_minor);
|
|
return 1;
|
|
}
|
|
int ptx_mb = size_mb;
|
|
if (!run_ptx_fallback(&cuda, dev, name, cc_major, cc_minor, seconds, ptx_mb, &report)) {
|
|
return 1;
|
|
}
|
|
}
|
|
|
|
print_stress_report(&report, device_index, seconds);
|
|
return 0;
|
|
}
|