Skip to content

Commit

Permalink
Fix improper clang format (#894)
Browse files Browse the repository at this point in the history
  • Loading branch information
jmachado-amd authored Feb 21, 2025
1 parent a0f6efc commit 5a20cb6
Show file tree
Hide file tree
Showing 3 changed files with 21 additions and 23 deletions.
19 changes: 9 additions & 10 deletions library/src/auxiliary/rocauxiliary_stedc.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -51,7 +51,6 @@ ROCSOLVER_BEGIN_NAMESPACE
// external gemm-based updates.
#define STEDC_EXTERNAL_GEMM true


typedef enum rocsolver_stedc_mode_
{
rocsolver_stedc_mode_qr,
Expand Down Expand Up @@ -1576,7 +1575,6 @@ ROCSOLVER_KERNEL void __launch_bounds__(STEDC_BDIM)
}
}


//--------------------------------------------------------------------------------------//
/** STEDC_MERGEVECTORS_KERNEL prepares vectors from the secular equation for
every pair of sub-blocks that need to be merged in a split block. A matrix in the batch
Expand Down Expand Up @@ -1799,7 +1797,8 @@ ROCSOLVER_KERNEL void __launch_bounds__(STEDC_BDIM)
if(idd[k] == 0)
dd++;
}
temps[pers[i - dd] + in + (p2 + j) * n] = vecs[i - dd - in + (p2 + j) * n] / nrm;
temps[pers[i - dd] + in + (p2 + j) * n]
= vecs[i - dd - in + (p2 + j) * n] / nrm;
}
else
temps[i + (p2 + j) * n] = 0;
Expand Down Expand Up @@ -1848,7 +1847,6 @@ ROCSOLVER_KERNEL void __launch_bounds__(STEDC_BDIM)
}
}


//--------------------------------------------------------------------------------------//
/** STEDC_MERGEUPDATE_KERNEL updates vectors after merges are done. A matrix in the batch
could have many split-blocks, and each split-block could be divided in a maximum of nn sub-blocks.
Expand Down Expand Up @@ -2426,10 +2424,11 @@ rocblas_status rocsolver_stedc_template(rocblas_handle handle,
E + shiftE, strideE, tmpz, tempgemm, splits, eps, ssfmin, ssfmax);

// c. find merged eigen vectors
ROCSOLVER_LAUNCH_KERNEL((stedc_mergeVectors_kernel<rocsolver_stedc_mode_qr, STEDC_EXTERNAL_GEMM, S>),
dim3(numgrps3, STEDC_NUM_SPLIT_BLKS, batch_count),
dim3(STEDC_BDIM), lmemsize3, stream, k, n, D + shiftD, strideD,
E + shiftE, strideE, V, 0, ldv, strideV, tmpz, tempgemm, splits);
ROCSOLVER_LAUNCH_KERNEL(
(stedc_mergeVectors_kernel<rocsolver_stedc_mode_qr, STEDC_EXTERNAL_GEMM, S>),
dim3(numgrps3, STEDC_NUM_SPLIT_BLKS, batch_count), dim3(STEDC_BDIM), lmemsize3,
stream, k, n, D + shiftD, strideD, E + shiftE, strideE, V, 0, ldv, strideV, tmpz,
tempgemm, splits);

if(STEDC_EXTERNAL_GEMM)
{
Expand All @@ -2439,8 +2438,8 @@ rocblas_status rocsolver_stedc_template(rocblas_handle handle,
// STEDC_EXTERNAL_GEMM at run time to switch between internal vector updates and
// external gemm based updates.
rocsolver_gemm(handle, rocblas_operation_none, rocblas_operation_none, n, n, n,
&one, V, 0, ldv, strideV, tempgemm, n*n, n, 2*n*n, &zero, tempgemm, 0, n, 2*n*n,
batch_count, workArr);
&one, V, 0, ldv, strideV, tempgemm, n * n, n, 2 * n * n, &zero,
tempgemm, 0, n, 2 * n * n, batch_count, workArr);
}

// d. update level
Expand Down
8 changes: 4 additions & 4 deletions library/src/auxiliary/rocauxiliary_stedcj.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -454,10 +454,10 @@ rocblas_status rocsolver_stedcj_template(rocblas_handle handle,
eps, ssfmin, ssfmax);

// c. find merged eigen vectors
ROCSOLVER_LAUNCH_KERNEL((stedc_mergeVectors_kernel<rocsolver_stedc_mode_jacobi, STEDCJ_EXTERNAL_GEMM, S>),
dim3(numgrps3, STEDC_NUM_SPLIT_BLKS, batch_count), dim3(STEDC_BDIM),
lmemsize3, stream, k, n, D, strideD, E, strideE, tempvect, 0, ldt,
strideT, tmpz, tempgemm, splits_map);
ROCSOLVER_LAUNCH_KERNEL(
(stedc_mergeVectors_kernel<rocsolver_stedc_mode_jacobi, STEDCJ_EXTERNAL_GEMM, S>),
dim3(numgrps3, STEDC_NUM_SPLIT_BLKS, batch_count), dim3(STEDC_BDIM), lmemsize3, stream,
k, n, D, strideD, E, strideE, tempvect, 0, ldt, strideT, tmpz, tempgemm, splits_map);

// c. update level
ROCSOLVER_LAUNCH_KERNEL((stedc_mergeUpdate_kernel<rocsolver_stedc_mode_jacobi, S>),
Expand Down
17 changes: 8 additions & 9 deletions library/src/auxiliary/rocauxiliary_stedcx.hpp
Original file line number Diff line number Diff line change
@@ -1,5 +1,5 @@
/************************************************************************
* Copyright (C) 2024 Advanced Micro Devices, Inc. All rights reserved.
* Copyright (C) 2024-2025 Advanced Micro Devices, Inc. All rights reserved.
*
* Redistribution and use in source and binary forms, with or without
* modification, are permitted provided that the following conditions
Expand Down Expand Up @@ -36,11 +36,10 @@

ROCSOLVER_BEGIN_NAMESPACE

// TODO: using macro STEDCX_EXTERNAL_GEMM = false for now. We can enable the use of
// external gemm updates once the development is completed for stedc.
// TODO: using macro STEDCX_EXTERNAL_GEMM = false for now. We can enable the use of
// external gemm updates once the development is completed for stedc.
#define STEDCX_EXTERNAL_GEMM false


/***************** Device auxiliary functions *****************************************/
/**************************************************************************************/

Expand Down Expand Up @@ -583,11 +582,11 @@ rocblas_status rocsolver_stedcx_template(rocblas_handle handle,
eps, ssfmin, ssfmax);

// c. find merged eigen vectors
ROCSOLVER_LAUNCH_KERNEL((stedc_mergeVectors_kernel<rocsolver_stedc_mode_bisection, STEDCX_EXTERNAL_GEMM, S>),
dim3(numgrps3, STEDC_NUM_SPLIT_BLKS, batch_count), dim3(STEDC_BDIM),
lmemsize3, stream, k, n, D, strideD, E, strideE, tempvect, 0, ldt,
strideT, tmpz, tempgemm, splits);
ROCSOLVER_LAUNCH_KERNEL(
(stedc_mergeVectors_kernel<rocsolver_stedc_mode_bisection, STEDCX_EXTERNAL_GEMM, S>),
dim3(numgrps3, STEDC_NUM_SPLIT_BLKS, batch_count), dim3(STEDC_BDIM), lmemsize3, stream,
k, n, D, strideD, E, strideE, tempvect, 0, ldt, strideT, tmpz, tempgemm, splits);

// d. update level
ROCSOLVER_LAUNCH_KERNEL((stedc_mergeUpdate_kernel<rocsolver_stedc_mode_bisection, S>),
dim3(numgrps3, STEDC_NUM_SPLIT_BLKS, batch_count), dim3(STEDC_BDIM),
Expand Down

0 comments on commit 5a20cb6

Please sign in to comment.