| 40 | |
| 41 | template<typename TA, typename TB> |
| 42 | void checkMat(TA* A, TB* B, int size, const char* mark, float threshold) |
| 43 | { |
| 44 | float max_diff = -10000.0f; |
| 45 | float max_diff_a, max_diff_b; |
| 46 | TA* matA = (TA*)malloc(sizeof(TA) * size); |
| 47 | TB* matB = (TB*)malloc(sizeof(TB) * size); |
| 48 | int not_passed = 0; |
| 49 | cudaMemcpy(matA, A, sizeof(TA) * size, cudaMemcpyDeviceToHost); |
| 50 | cudaMemcpy(matB, B, sizeof(TB) * size, cudaMemcpyDeviceToHost); |
| 51 | float A_nonZero_ratio = float(checkNonZero(A, size)) / float(size); |
| 52 | float B_nonZero_ratio = float(checkNonZero(B, size)) / float(size); |
| 53 | if (A_nonZero_ratio < 0.1 || B_nonZero_ratio < 0.1) { |
| 54 | printf("[%s] nonZero ratio [%f] [%f]\n", mark, A_nonZero_ratio, B_nonZero_ratio); |
| 55 | } |
| 56 | for (int jjj = 0; jjj < size; jjj++) { |
| 57 | float diff = fabs(float(matA[jjj]) - float(matB[jjj])); |
| 58 | if (diff > max_diff) { |
| 59 | max_diff = diff; |
| 60 | max_diff_a = float(matA[jjj]); |
| 61 | max_diff_b = float(matB[jjj]); |
| 62 | } |
| 63 | if (fabs(float(matA[jjj]) - float(matB[jjj])) > threshold) { |
| 64 | not_passed += 1; |
| 65 | if (not_passed < 1000) { |
| 66 | printf("%d %f %f %f\n", jjj, float(matA[jjj]), float(matB[jjj]), float(matA[jjj]) - float(matB[jjj])); |
| 67 | } |
| 68 | } |
| 69 | } |
| 70 | FT_LOG_INFO("[%s] max diff : %f ; a : %f ; b : %f", mark, max_diff, max_diff_a, max_diff_b); |
| 71 | if (not_passed != 0) { |
| 72 | FT_LOG_ERROR("[%s] different elements : %d ", mark, not_passed); |
| 73 | FT_CHECK(false); |
| 74 | } |
| 75 | else { |
| 76 | FT_LOG_INFO("[%s] check pass!", mark); |
| 77 | } |
| 78 | free(matA); |
| 79 | free(matB); |
| 80 | } |
| 81 | |
| 82 | template<typename T> |
| 83 | void layernorm_test(const int m, const int n); |
no test coverage detected